Replace and deprecate thrust::cuda_cub::terminate (#3421) (#3425)

NVIDIA · Jan 20, 2025 · 406d22d · 406d22d
1 parent c942552
commit 406d22d
Show file tree

Hide file tree

Showing 2 changed files with 6 additions and 12 deletions.
diff --git a/thrust/thrust/system/cuda/detail/terminate.h b/thrust/thrust/system/cuda/detail/terminate.h
@@ -36,7 +36,8 @@
 #elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC)
 #  pragma system_header
 #endif // no system header
-#include <thrust/system/cuda/detail/util.h>
+
+#include <cuda/std/__exception/terminate.h>
 
 #include <cstdio>
 
@@ -47,18 +48,11 @@ namespace cuda
 {
 namespace detail
 {
-
-inline _CCCL_DEVICE void terminate()
-{
-  thrust::cuda_cub::terminate();
-}
-
 inline _CCCL_HOST_DEVICE void terminate_with_message(const char* message)
 {
   printf("%s\n", message);
-  thrust::cuda_cub::terminate();
+  ::cuda::std::terminate();
 }
-
 } // namespace detail
 } // namespace cuda
 } // namespace system

diff --git a/thrust/thrust/system/cuda/detail/util.h b/thrust/thrust/system/cuda/detail/util.h
@@ -178,7 +178,7 @@ trivial_copy_device_to_device(Policy& policy, Type* dst, Type const* src, size_t
   return status;
 }
 
-inline void _CCCL_HOST_DEVICE terminate()
+CCCL_DEPRECATED_BECAUSE("Use cuda::std::terminate() instead") inline void _CCCL_HOST_DEVICE terminate()
 {
   NV_IF_TARGET(NV_IS_HOST, (std::terminate();), (asm("trap;");));
 }
@@ -210,7 +210,7 @@ _CCCL_HOST_DEVICE inline void throw_on_error(cudaError_t status)
 
     NV_IF_TARGET(NV_IS_HOST,
                  (throw thrust::system_error(status, thrust::cuda_category());),
-                 (THRUST_TEMP_DEVICE_CODE; cuda_cub::terminate();));
+                 (THRUST_TEMP_DEVICE_CODE; ::cuda::std::terminate();));
 
 #undef THRUST_TEMP_DEVICE_CODE
   }
@@ -243,7 +243,7 @@ _CCCL_HOST_DEVICE inline void throw_on_error(cudaError_t status, char const* msg
 
     NV_IF_TARGET(NV_IS_HOST,
                  (throw thrust::system_error(status, thrust::cuda_category(), msg);),
-                 (THRUST_TEMP_DEVICE_CODE; cuda_cub::terminate();));
+                 (THRUST_TEMP_DEVICE_CODE; ::cuda::std::terminate();));
 
 #undef THRUST_TEMP_DEVICE_CODE
   }