General: Bump HIP backend to ROCm 7.1+

cmake/hip/Findthrust.cmake

@@ -0,0 +1,32 @@
+find_package(rocthrust)
+
+if(rocthrust_INCLUDE_DIR)
+    file(STRINGS "${rocthrust_INCLUDE_DIR}/thrust/version.h"
+         THRUST_VERSION_STRING
+         REGEX "#define THRUST_VERSION[ \t]+([0-9x]+)")
+
+    string(REGEX REPLACE "#define THRUST_VERSION[ \t]+([0-9]+).*" "\\1" THRUST_VERSION_STRING ${THRUST_VERSION_STRING})
+
+    math(EXPR THRUST_VERSION_MAJOR "${THRUST_VERSION_STRING} / 100000")
+    math(EXPR THRUST_VERSION_MINOR "(${THRUST_VERSION_STRING} / 100) % 1000")
+    math(EXPR THRUST_VERSION_PATCH "${THRUST_VERSION_STRING} % 100")
+    unset(THRUST_VERSION_STRING)
+
+    set(THRUST_VERSION "${THRUST_VERSION_MAJOR}.${THRUST_VERSION_MINOR}.${THRUST_VERSION_PATCH}")
+endif()
+
+include(FindPackageHandleStandardArgs)
+find_package_handle_standard_args(thrust
+                                REQUIRED_VARS rocthrust_INCLUDE_DIR
+                                VERSION_VAR THRUST_VERSION)
+
+if(thrust_FOUND)
+    add_library(thrust::thrust INTERFACE IMPORTED)
+    target_link_libraries(thrust::thrust INTERFACE roc::rocthrust)
+
+    mark_as_advanced(THRUST_INCLUDE_DIR
+                     THRUST_VERSION
+                     THRUST_VERSION_MAJOR
+                     THRUST_VERSION_MINOR
+                     THRUST_VERSION_PATCH)
+endif()

cmake/hip/set_device_flags.cmake

@@ -21,10 +21,6 @@ function(stdgpu_set_device_flags STDGPU_OUTPUT_DEVICE_FLAGS)
         endif()
     endif()
 
-    if(CMAKE_HIP_COMPILER_ID STREQUAL "Clang")
-        list(APPEND ${STDGPU_OUTPUT_DEVICE_FLAGS} "-Wno-pass-failed")
-    endif()
-
     set(${STDGPU_OUTPUT_DEVICE_FLAGS} "$<$<COMPILE_LANGUAGE:HIP>:${${STDGPU_OUTPUT_DEVICE_FLAGS}}>")
 
     # Make output variable visible

docs/CMakeLists.txt

@@ -30,7 +30,7 @@ endif()
 # - The dark mode switch and the respective variables of the sphinx-book-theme will be used
 #
 # Therefore, use the modified stylesheet from ROCm released under the MIT license:
-# https://github.com/RadeonOpenCompute/rocm-docs-core/blob/develop/src/rocm_docs/data/_doxygen/extra_stylesheet.css
+# https://github.com/ROCm/rocm-docs-core/blob/develop/src/rocm_docs/data/_doxygen/extra_stylesheet.css
 
 
 set(STDGPU_INCLUDE_DIR "${CMAKE_CURRENT_SOURCE_DIR}/../src")

docs/getting_started/building_from_source.md

@@ -48,7 +48,7 @@ Before building the library, please make sure that all required development tool
     - Clang 10: `sudo apt install clang`
 - HIP compiler
     - Clang (Already included in ROCm)
-- ROCm 5.1 <https://github.com/RadeonOpenCompute/ROCm>
+- ROCm 7.1 <https://github.com/ROCm/ROCm>
 - CMake 3.21.3: `sudo apt install cmake`
 
 :::
@@ -91,7 +91,7 @@ Before building the library, please make sure that all required development tool
     - MSVC 19.20 (Visual Studio 2019) <https://visualstudio.microsoft.com/downloads/>
 - HIP compiler
     - Clang (Already included in ROCm)
-- ROCm 5.1 <https://github.com/RadeonOpenCompute/ROCm>
+- ROCm 7.1 <https://github.com/ROCm/ROCm>
 - CMake 3.21.3: <https://cmake.org/download>
 
 :::

src/stdgpu/hip/CMakeLists.txt

@@ -1,8 +1,8 @@
-find_package(hip 5.1 REQUIRED)
+#find_package(hip 7.1 REQUIRED)
 
-set(STDGPU_DEPENDENCIES_BACKEND_INIT "
-find_dependency(hip 5.1 REQUIRED)
-" PARENT_SCOPE)
+#set(STDGPU_DEPENDENCIES_BACKEND_INIT "
+#find_dependency(hip 7.1 REQUIRED)
+#" PARENT_SCOPE)
 
 target_sources(stdgpu PRIVATE impl/device.cpp
                               impl/memory.cpp)
@@ -29,10 +29,11 @@ target_compile_features(stdgpu PUBLIC hip_std_17)
 
 target_compile_definitions(stdgpu PUBLIC THRUST_DEVICE_SYSTEM=THRUST_DEVICE_SYSTEM_HIP)
 
-# CMake always links libraries and executables involving HIP with the device linker, even if only host code is used.
-# Suppress the linker warning about unused HIP architectures *globally* via a public linker flag.
-set(STDGPU_HIP_DEVICE_LINKER_FLAGS "$<$<LINK_LANGUAGE:HIP>:-Wno-unused-command-line-argument>")
-message(STATUS "Created HIP device linker flags : ${STDGPU_HIP_DEVICE_LINKER_FLAGS}")
-target_link_options(stdgpu PUBLIC "${STDGPU_HIP_DEVICE_LINKER_FLAGS}")
+# thrust (rocthrust) already links against hip::device, which in turn links against hip::host
+#target_link_libraries(stdgpu PUBLIC hip::host)
 
-target_link_libraries(stdgpu PUBLIC hip::host)
+
+# Install custom thrust module
+install(FILES "${stdgpu_SOURCE_DIR}/cmake/${STDGPU_BACKEND_DIRECTORY}/Findthrust.cmake"
+        DESTINATION "${STDGPU_CMAKE_INSTALL_DIR}/${STDGPU_BACKEND_DIRECTORY}"
+        COMPONENT stdgpu)

src/stdgpu/hip/impl/memory_detail.h

@@ -33,8 +33,7 @@ memcpy_impl(ExecutionPolicy&& policy,
             hipMemcpyKind kind,
             bool needs_sychronization)
 {
-    cudaStream_t stream =
-            thrust::hip_rocprim::stream(thrust::detail::derived_cast(thrust::detail::strip_const(policy)));
+    hipStream_t stream = thrust::hip_rocprim::stream(thrust::detail::derived_cast(thrust::detail::strip_const(policy)));
 
     STDGPU_HIP_SAFE_CALL(hipMemcpyAsync(destination, source, static_cast<std::size_t>(bytes), kind, stream));
     if (needs_sychronization)

tools/backend/configure_hip.sh

@@ -12,4 +12,4 @@ fi
 sh tools/backend/helper/create_empty_directory.sh build
 
 # Configure project
-sh tools/backend/helper/configure.sh $CONFIG -DSTDGPU_BACKEND=STDGPU_BACKEND_HIP -DSTDGPU_COMPILE_WARNING_AS_ERROR=ON -DCMAKE_VERIFY_INTERFACE_HEADER_SETS=ON $@
+sh tools/backend/helper/configure.sh $CONFIG -DSTDGPU_BACKEND=STDGPU_BACKEND_HIP -DSTDGPU_COMPILE_WARNING_AS_ERROR=ON -DCMAKE_VERIFY_INTERFACE_HEADER_SETS=ON -DCMAKE_CXX_COMPILER=hipcc $@