Move hipification into dedicated files

mconcas · Sep 11, 2023 · 2641f42 · 2641f42
1 parent d454f03
commit 2641f42
Show file tree

Hide file tree

Showing 7 changed files with 186 additions and 69 deletions.
diff --git a/CMakeLists.txt b/CMakeLists.txt
@@ -93,6 +93,8 @@ include(O2DataFile)
 include(O2TargetManPage)
 include(O2AddWorkflow)
 include(O2SetROOTPCMDependencies)
+include(O2AddHipifiedExecutable)
+include (O2AddHipifiedLibrary)
 
 # Main targets of the project in various subdirectories. Order matters.
 add_subdirectory(Common)

diff --git a/Detectors/ITSMFT/ITS/tracking/GPU/hip/CMakeLists.txt b/Detectors/ITSMFT/ITS/tracking/GPU/hip/CMakeLists.txt
@@ -10,43 +10,43 @@
 # or submit itself to any jurisdiction.
 
 if(HIP_ENABLED)
-  # Hipify-perl to generate HIP sources
-  set(HIPIFY_EXECUTABLE "/opt/rocm/bin/hipify-perl")
-  file(GLOB CUDA_SOURCES_FULL_PATH "../cuda/*.cu")
-  foreach(file ${CUDA_SOURCES_FULL_PATH})
-    set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS ${file})
-    get_filename_component(CUDA_SOURCE ${file} NAME)
-    string(REPLACE ".cu" "" CUDA_SOURCE_NAME ${CUDA_SOURCE})
-    add_custom_command(
-      OUTPUT ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
-      COMMAND ${HIPIFY_EXECUTABLE} --quiet-warnings ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE} | sed '1{/\#include \"hip\\/hip_runtime.h\"/d}' > ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
-      DEPENDS ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE}
-    )
-  endforeach()
+  # # Hipify-perl to generate HIP sources
+  # set(HIPIFY_EXECUTABLE "/opt/rocm/bin/hipify-perl")
+  # file(GLOB CUDA_SOURCES_FULL_PATH "../cuda/*.cu")
+  # foreach(file ${CUDA_SOURCES_FULL_PATH})
+  #   set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS ${file})
+  #   get_filename_component(CUDA_SOURCE ${file} NAME)
+  #   string(REPLACE ".cu" "" CUDA_SOURCE_NAME ${CUDA_SOURCE})
+  #   add_custom_command(
+  #     OUTPUT ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
+  #     COMMAND ${HIPIFY_EXECUTABLE} --quiet-warnings ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE} | sed '1{/\#include \"hip\\/hip_runtime.h\"/d}' > ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
+  #     DEPENDS ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE}
+  #   )
+  # endforeach()
 
-  install(DIRECTORY ${CMAKE_CURRENT_SOURCE_DIR}/${baseTargetName}
-          DESTINATION ${CMAKE_INSTALL_INCLUDEDIR})
+  # install(DIRECTORY ${CMAKE_CURRENT_SOURCE_DIR}/${baseTargetName}
+  #         DESTINATION ${CMAKE_INSTALL_INCLUDEDIR})
 
   set(CMAKE_HIP_FLAGS "${CMAKE_HIP_FLAGS} ${O2_HIP_CMAKE_CXX_FLAGS} -fgpu-rdc")
   message(STATUS "Building ITS HIP tracker")
-  o2_add_library(ITStrackingHIP
-                 SOURCES ClusterLinesGPU.hip
-                         Context.hip
-                         TimeFrameGPU.hip
-                         Stream.hip
-                         TrackerTraitsGPU.hip
-                         TracerGPU.hip
-                         VertexerTraitsGPU.hip
-                         Utils.hip
+  o2_add_hipified_library(ITStrackingHIP
+                 SOURCES ../cuda/ClusterLinesGPU.cu
+                         ../cuda/Context.cu
+                         ../cuda/TimeFrameGPU.cu
+                         ../cuda/Stream.cu
+                         ../cuda/TrackerTraitsGPU.cu
+                         ../cuda/TracerGPU.cu
+                         ../cuda/VertexerTraitsGPU.cu
+                         ../cuda/Utils.cu
                  PUBLIC_INCLUDE_DIRECTORIES ../
                  PUBLIC_LINK_LIBRARIES O2::ITStracking
                                        hip::host
                                        hip::device
                                        hip::hipcub
                  TARGETVARNAME targetName)
 
-  target_compile_definitions(
-    ${targetName} PRIVATE $<TARGET_PROPERTY:O2::ITStracking,COMPILE_DEFINITIONS>)
+  # target_compile_definitions(
+  #   ${targetName} PRIVATE $<TARGET_PROPERTY:O2::ITStracking,COMPILE_DEFINITIONS>)
 
   if(O2_HIP_CMAKE_LINK_FLAGS)
     # Need to add gpu target also to link flags due to gpu-rdc option

diff --git a/GPU/GPUbenchmark/Shared/Utils.h b/GPU/GPUbenchmark/Shared/Utils.h
@@ -47,6 +47,11 @@
   exit(EXIT_FAILURE);
 #endif
 
+template <typename T>
+void discardResult(const T&)
+{
+}
+
 enum class Test {
   Read,
   Write,

diff --git a/GPU/GPUbenchmark/cuda/Kernels.cu b/GPU/GPUbenchmark/cuda/Kernels.cu
@@ -428,16 +428,14 @@ float GPUbenchmark<chunk_t>::runSequential(void (*kernel)(chunk_t*, size_t, T...
 
   // Warm up
   (*kernel)<<<nBlocks, nThreads, 0, stream>>>(chunkPtr, getBufferCapacity<chunk_t>(chunk.second, mOptions.prime), args...);
-  cudaDeviceSynchronize();
   GPUCHECK(cudaGetLastError());
   GPUCHECK(cudaEventCreate(&start));
   GPUCHECK(cudaEventCreate(&stop));
 
   GPUCHECK(cudaEventRecord(start));
   for (auto iLaunch{0}; iLaunch < nLaunches; ++iLaunch) {                                                                     // Schedule all the requested kernel launches
     (*kernel)<<<nBlocks, nThreads, 0, stream>>>(chunkPtr, getBufferCapacity<chunk_t>(chunk.second, mOptions.prime), args...); // NOLINT: clang-tidy false-positive
-    cudaDeviceSynchronize();
-    GPUCHECK(cudaGetLastError());
+      GPUCHECK(cudaGetLastError());
   }
   GPUCHECK(cudaEventRecord(stop));      // record checkpoint
   GPUCHECK(cudaEventSynchronize(stop)); // synchronize executions

diff --git a/GPU/GPUbenchmark/hip/CMakeLists.txt b/GPU/GPUbenchmark/hip/CMakeLists.txt
@@ -10,43 +10,8 @@
 # or submit itself to any jurisdiction.
 
 message(STATUS "Building GPU HIP benchmark")
-# Hipify-perl to generate HIP sources
-set(HIPIFY_EXECUTABLE "/opt/rocm/bin/hipify-perl")
-file(GLOB CUDA_SOURCES_FULL_PATH "../cuda/*.cu")
-foreach(file ${CUDA_SOURCES_FULL_PATH})
-  set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS ${file})
-  get_filename_component(CUDA_SOURCE ${file} NAME)
-  string(REPLACE ".cu" "" CUDA_SOURCE_NAME ${CUDA_SOURCE})
-  add_custom_command(
-    OUTPUT ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
-    COMMAND ${HIPIFY_EXECUTABLE} --quiet-warnings ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE} | sed '1{/\#include \"hip\\/hip_runtime.h\"/d}' > ${CMAKE_CURRENT_SOURCE_DIR}/${CUDA_SOURCE_NAME}.hip
-    DEPENDS ${CMAKE_CURRENT_SOURCE_DIR}/../cuda/${CUDA_SOURCE}
-  )
-endforeach()
-message(STATUS "CMAKE_EXE_LINKER_FLAGS: ${CMAKE_EXE_LINKER_FLAGS}")
-message(STATUS "CMAKE_SHARED_LINKER_FLAGS: ${CMAKE_SHARED_LINKER_FLAGS}")
-message(STATUS "CMAKE_HIP_COMPILE_FEATURES: ${CMAKE_HIP_COMPILE_FEATURES}")
-message(STATUS "CMAKE_LINKER: ${CMAKE_LINKER}")
-message(STATUS "CMAKE_HIP_STANDARD_LIBRARIES: ${CMAKE_HIP_STANDARD_LIBRARIES}")
-message(STATUS "CMAKE_HIP_LINK_EXECUTABLE: ${CMAKE_HIP_LINK_EXECUTABLE}")
-message(STATUS "CMAKE_HIP_LINK_FLAGS: ${CMAKE_HIP_LINK_FLAGS}")
-message(STATUS "CMAKE_HIP_RUNTIME_LIBRARIES_STATIC: ${CMAKE_HIP_RUNTIME_LIBRARIES_STATIC}")
-message(STATUS "CMAKE_HIP_RUNTIME_LIBRARIES_SHARED: ${CMAKE_HIP_RUNTIME_LIBRARIES_SHARED}")
-message(STATUS "CMAKE_MODULE_LINKER_FLAGS: ${CMAKE_MODULE_LINKER_FLAGS}")
-o2_add_executable(gpu-memory-benchmark-hip
-                  SOURCES benchmark.hip
-                          Kernels.hip
-                  PUBLIC_LINK_LIBRARIES hip::host
-                                        hip-lang::device
-                                        Boost::program_options
-                  TARGETVARNAME targetName)
-# >:-)
-include_directories(${Boost_INCLUDE_DIRS})
-add_executable(alternative-gpu-memory-benchmark-hip benchmark.hip Kernels.hip)
-target_link_libraries(alternative-gpu-memory-benchmark-hip Boost::program_options ROOT::Tree)
-install(TARGETS alternative-gpu-memory-benchmark-hip RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR})
-
-# if(O2_HIP_CMAKE_LINK_FLAGS)
-  # Need to add gpu target also to link flags due to gpu-rdc option
-  # target_link_options(${targetName} PUBLIC ${O2_HIP_CMAKE_LINK_FLAGS})
-# endif()
+o2_add_hipified_executable(gpu-memory-benchmark-hip
+                           SOURCES ../cuda/benchmark.cu
+                                   ../cuda/Kernels.cu
+                           PUBLIC_LINK_LIBRARIES hip::host
+                                                 Boost::program_options)
diff --git a/cmake/O2AddHipifiedExecutable.cmake b/cmake/O2AddHipifiedExecutable.cmake
@@ -0,0 +1,76 @@
+# Copyright 2019-2020 CERN and copyright holders of ALICE O2.
+# See https://alice-o2.web.cern.ch/copyright for details of the copyright holders.
+# All rights not expressly granted are reserved.
+#
+# This software is distributed under the terms of the GNU General Public
+# License v3 (GPL Version 3), copied verbatim in the file "COPYING".
+#
+# In applying this license CERN does not waive the privileges and immunities
+# granted to it by virtue of its status as an Intergovernmental Organization
+# or submit itself to any jurisdiction.
+
+include_guard()
+
+include(O2AddExecutable)
+
+function(o2_add_hipified_executable baseTargetName)
+  # Parse arguments in the same way o2_add_executable does
+  cmake_parse_arguments(PARSE_ARGV
+                        1
+                        A
+                        "IS_TEST;NO_INSTALL;IS_BENCHMARK"
+                        "COMPONENT_NAME;TARGETVARNAME"
+                        "SOURCES;PUBLIC_LINK_LIBRARIES;JOB_POOL")
+
+  # Process each .cu file to generate a .hip file
+  set(HIPIFY_EXECUTABLE "/opt/rocm/bin/hipify-perl")
+  set(HIP_SOURCES)
+
+  foreach(file ${A_SOURCES})
+    get_filename_component(ABS_CUDA_SORUCE ${file} ABSOLUTE)
+    if(file MATCHES "\\.cu$")
+      set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS ${file})
+      get_filename_component(CUDA_SOURCE ${file} NAME)
+      string(REPLACE ".cu" ".hip" HIP_SOURCE ${CUDA_SOURCE})
+      set(OUTPUT_HIP_FILE "${CMAKE_CURRENT_SOURCE_DIR}/${HIP_SOURCE}")
+      list(APPEND HIP_SOURCES ${OUTPUT_HIP_FILE})
+
+      add_custom_command(
+        OUTPUT ${OUTPUT_HIP_FILE}
+        COMMAND ${HIPIFY_EXECUTABLE} --quiet-warnings ${ABS_CUDA_SORUCE} | sed '1{/\#include \"hip\\/hip_runtime.h\"/d}' > ${OUTPUT_HIP_FILE}
+        DEPENDS ${file}
+      )
+    else()
+      list(APPEND HIP_SOURCES ${file})
+    endif()
+  endforeach()
+
+  # This is a bit cumbersome, but it seems the only suitable since cmake_parse_arguments is not capable to filter only the SOURCE variadic values 
+  set(FORWARD_ARGS "")
+  if(A_IS_TEST)
+    list(APPEND FORWARD_ARGS "IS_TEST")
+  endif()
+  if(A_NO_INSTALL)
+    list(APPEND FORWARD_ARGS "NO_INSTALL")
+  endif()
+  if(A_IS_BENCHMARK)
+    list(APPEND FORWARD_ARGS "IS_BENCHMARK")
+  endif()
+  if(A_COMPONENT_NAME)
+    list(APPEND FORWARD_ARGS "COMPONENT_NAME" ${A_COMPONENT_NAME})
+  endif()
+  if(A_TARGETVARNAME)
+    list(APPEND FORWARD_ARGS "TARGETVARNAME" ${A_TARGETVARNAME})
+  endif()
+  if(A_PUBLIC_LINK_LIBRARIES)
+    list(APPEND FORWARD_ARGS "PUBLIC_LINK_LIBRARIES" ${A_PUBLIC_LINK_LIBRARIES})
+  endif()
+  if(A_JOB_POOL)
+    list(APPEND FORWARD_ARGS "JOB_POOL" ${A_JOB_POOL})
+  endif()
+
+  # Call o2_add_executable with new sources
+  o2_add_executable("${baseTargetName}"
+                    SOURCES ${HIP_SOURCES}
+                    ${FORWARD_ARGS})
+endfunction()
diff --git a/cmake/O2AddHipifiedLibrary.cmake b/cmake/O2AddHipifiedLibrary.cmake
@@ -0,0 +1,71 @@
+# Copyright 2019-2020 CERN and copyright holders of ALICE O2.
+# See https://alice-o2.web.cern.ch/copyright for details of the copyright holders.
+# All rights not expressly granted are reserved.
+#
+# This software is distributed under the terms of the GNU General Public
+# License v3 (GPL Version 3), copied verbatim in the file "COPYING".
+#
+# In applying this license CERN does not waive the privileges and immunities
+# granted to it by virtue of its status as an Intergovernmental Organization
+# or submit itself to any jurisdiction.
+
+include_guard()
+
+include(O2AddLibrary)
+
+function(o2_add_hipified_library baseTargetName)
+  # Parse arguments in the same way o2_add_library does
+  cmake_parse_arguments(PARSE_ARGV
+                        1
+                        A
+                        ""
+                        "TARGETVARNAME"
+                        "SOURCES;PUBLIC_INCLUDE_DIRECTORIES;PUBLIC_LINK_LIBRARIES;PRIVATE_INCLUDE_DIRECTORIES;PRIVATE_LINK_LIBRARIES"
+                        )
+
+  # Process each .cu file to generate a .hip file
+  set(HIPIFY_EXECUTABLE "/opt/rocm/bin/hipify-perl")
+  set(HIP_SOURCES)
+
+  foreach(file ${A_SOURCES})
+    get_filename_component(ABS_CUDA_SORUCE ${file} ABSOLUTE)
+    if(file MATCHES "\\.cu$")
+      set_property(DIRECTORY APPEND PROPERTY CMAKE_CONFIGURE_DEPENDS ${file})
+      get_filename_component(CUDA_SOURCE ${file} NAME)
+      string(REPLACE ".cu" ".hip" HIP_SOURCE ${CUDA_SOURCE})
+      set(OUTPUT_HIP_FILE "${CMAKE_CURRENT_SOURCE_DIR}/${HIP_SOURCE}")
+      list(APPEND HIP_SOURCES ${OUTPUT_HIP_FILE})
+
+      add_custom_command(
+        OUTPUT ${OUTPUT_HIP_FILE}
+        COMMAND ${HIPIFY_EXECUTABLE} --quiet-warnings ${ABS_CUDA_SORUCE} | sed '1{/\#include \"hip\\/hip_runtime.h\"/d}' > ${OUTPUT_HIP_FILE}
+        DEPENDS ${file}
+      )
+    else()
+      list(APPEND HIP_SOURCES ${file})
+    endif()
+  endforeach()
+
+  # This is a bit cumbersome, but it seems the only suitable since cmake_parse_arguments is not capable to filter only the SOURCE variadic values 
+  set(FORWARD_ARGS "")
+  if(A_TARGETVARNAME)
+    list(APPEND FORWARD_ARGS "TARGETVARNAME" ${A_TARGETVARNAME})
+  endif()
+  if(A_PUBLIC_INCLUDE_DIRECTORIES)
+    list(APPEND FORWARD_ARGS "PUBLIC_INCLUDE_DIRECTORIES" ${A_PUBLIC_INCLUDE_DIRECTORIES})
+  endif()
+  if(A_PUBLIC_LINK_LIBRARIES)
+    list(APPEND FORWARD_ARGS "PUBLIC_LINK_LIBRARIES" ${A_PUBLIC_LINK_LIBRARIES})
+  endif()
+  if(A_PRIVATE_INCLUDE_DIRECTORIES)
+    list(APPEND FORWARD_ARGS "PRIVATE_INCLUDE_DIRECTORIES" ${A_PRIVATE_INCLUDE_DIRECTORIES})
+  endif()
+  if(A_PRIVATE_LINK_LIBRARIES)
+    list(APPEND FORWARD_ARGS "PRIVATE_LINK_LIBRARIES" ${A_PRIVATE_LINK_LIBRARIES})
+  endif()
+
+  # Call o2_add_library with new sources
+  o2_add_library("${baseTargetName}"
+                 SOURCES ${HIP_SOURCES}
+                 ${FORWARD_ARGS})
+endfunction()