Merge branch 'main' into fea/build-rapids

NVIDIA · May 3, 2024 · cc20473 · cc20473
2 parents 0928bac + c910b16
commit cc20473
Show file tree

Hide file tree

Showing 5 changed files with 56 additions and 11 deletions.
diff --git a/CMakePresets.json b/CMakePresets.json
@@ -56,6 +56,16 @@
         "THRUST_MULTICONFIG_ENABLE_SYSTEM_TBB": true
       }
     },
+    {
+      "name": "all-dev-debug",
+      "displayName": "all-dev debug",
+      "inherits": "all-dev",
+      "cacheVariables":{
+        "CCCL_ENABLE_BENCHMARKS": false,
+        "CMAKE_BUILD_TYPE": "Debug",
+        "CMAKE_CUDA_FLAGS": "-G"
+      }
+    },
     {
       "name": "libcudacxx-codegen",
       "displayName": "libcu++: codegen",
@@ -271,23 +281,17 @@
         "CCCL_ENABLE_EXAMPLES": true,
         "CCCL_ENABLE_TESTING": true
       }
-    },
-    {
-      "name": "all-dev-debug",
-      "displayName": "all-dev debug",
-      "inherits": "all-dev",
-      "cacheVariables":{
-        "CCCL_ENABLE_BENCHMARKS": false,
-        "CMAKE_BUILD_TYPE": "Debug",
-        "CMAKE_CUDA_FLAGS": "-G"
-      }
     }
   ],
   "buildPresets": [
     {
       "name": "all-dev",
       "configurePreset": "all-dev"
     },
+    {
+      "name": "all-dev-debug",
+      "configurePreset": "all-dev-debug"
+    },
     {
       "name": "libcudacxx-codegen",
       "configurePreset": "libcudacxx-codegen",
@@ -411,6 +415,11 @@
       "configurePreset": "all-dev",
       "inherits": "base"
     },
+    {
+      "name": "all-dev-debug",
+      "configurePreset": "all-dev-debug",
+      "inherits": "all-dev"
+    },
     {
       "name": "libcudacxx-ctest-base",
       "hidden": true,

diff --git a/cub/cmake/CubInstallRules.cmake b/cub/cmake/CubInstallRules.cmake
@@ -8,6 +8,7 @@ install(DIRECTORY "${CUB_SOURCE_DIR}/cub"
   DESTINATION "${CMAKE_INSTALL_INCLUDEDIR}"
   FILES_MATCHING
     PATTERN "*.cuh"
+    PATTERN "*.hpp"
 )
 
 install(DIRECTORY "${CUB_SOURCE_DIR}/cub/cmake/"

diff --git a/cub/test/CMakeLists.txt b/cub/test/CMakeLists.txt
@@ -26,6 +26,15 @@ option(METAL_BUILD_EXAMPLES OFF)
 option(METAL_BUILD_TESTS OFF)
 CPMAddPackage("gh:brunocodutra/[email protected]")
 
+CPMAddPackage(
+  NAME NVTX
+  GITHUB_REPOSITORY NVIDIA/NVTX
+  GIT_TAG release-v3
+  DOWNLOAD_ONLY
+  SYSTEM
+)
+include("${NVTX_SOURCE_DIR}/c/nvtxImportedTargets.cmake")
+
 find_package(CUDAToolkit)
 
 set(curand_default OFF)
@@ -280,6 +289,10 @@ function(cub_add_test target_name_var test_name test_src cub_target launcher_id)
     target_include_directories(${test_target} PRIVATE "${CUB_SOURCE_DIR}/test")
     target_compile_definitions(${test_target} PRIVATE CUB_DETAIL_DEBUG_ENABLE_SYNC)
 
+    if ("${test_target}" MATCHES "nvtx_in_usercode")
+      target_link_libraries(${test_target} nvtx3-cpp)
+    endif()
+
     if (CUB_IN_THRUST)
       thrust_fix_clang_nvcc_build_for(${test_target})
     endif()

diff --git a/cub/test/catch2_segmented_sort_helper.cuh b/cub/test/catch2_segmented_sort_helper.cuh
@@ -646,7 +646,7 @@ CUB_RUNTIME_FUNCTION cudaError_t call_cub_segmented_sort_api(
   const int* offset_begin_it                  = d_begin_offsets;
   thrust::device_ptr<const int> offset_end_it = thrust::device_pointer_cast(d_end_offsets);
 
-  cudaError_t status = cudaErrorNotYetImplemented;
+  cudaError_t status = cudaErrorInvalidValue;
 
   if (stable_sort)
   {

diff --git a/cub/test/test_nvtx_in_usercode.cu b/cub/test/test_nvtx_in_usercode.cu
@@ -0,0 +1,22 @@
+#include <cub/device/device_for.cuh> // internal include of NVTX
+
+#include <thrust/iterator/counting_iterator.h>
+
+#include <nvtx3/nvtx3.hpp> // user-side include of NVTX, retrieved elsewhere
+
+struct Op
+{
+  _CCCL_HOST_DEVICE void operator()(int i) const
+  {
+    printf("%d\n", i);
+  }
+};
+
+int main()
+{
+  nvtx3::scoped_range range("user-range"); // user-side use of NVTX
+
+  thrust::counting_iterator<int> it{0};
+  cub::DeviceFor::ForEach(it, it + 16, Op{}); // internal use of NVTX
+  cudaDeviceSynchronize();
+}