Deprecate thrust::async

Fixes: #100
NVIDIA · Jan 10, 2025 · 80bef1d · 80bef1d
1 parent 38e3d0d
commit 80bef1d
Show file tree

Hide file tree

Showing 12 changed files with 103 additions and 53 deletions.
diff --git a/thrust/testing/async_copy.cu b/thrust/testing/async_copy.cu
@@ -10,6 +10,8 @@
 #  include <unittest/unittest.h>
 #  include <unittest/util_async.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 #  define DEFINE_ASYNC_COPY_CALLABLE(name, ...)                                               \
     struct THRUST_PP_CAT2(name, _fn)                                                          \
     {                                                                                         \

diff --git a/thrust/testing/async_for_each.cu b/thrust/testing/async_for_each.cu
@@ -8,6 +8,8 @@
 
 #  include <unittest/unittest.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 #  define DEFINE_ASYNC_FOR_EACH_CALLABLE(name, ...)                                           \
     struct THRUST_PP_CAT2(name, _fn)                                                          \
     {                                                                                         \
@@ -21,7 +23,6 @@
     /**/
 
 DEFINE_ASYNC_FOR_EACH_CALLABLE(invoke_async_for_each);
-
 DEFINE_ASYNC_FOR_EACH_CALLABLE(invoke_async_for_each_device, thrust::device);
 
 #  undef DEFINE_ASYNC_FOR_EACH_CALLABLE

diff --git a/thrust/testing/async_reduce.cu b/thrust/testing/async_reduce.cu
@@ -12,6 +12,8 @@
 #  include <unittest/unittest.h>
 #  include <unittest/util_async.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 template <typename T>
 struct custom_plus
 {

diff --git a/thrust/testing/async_reduce_into.cu b/thrust/testing/async_reduce_into.cu
@@ -13,6 +13,8 @@
 #  include <unittest/unittest.h>
 #  include <unittest/util_async.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 template <typename T>
 struct custom_plus
 {

diff --git a/thrust/testing/async_sort.cu b/thrust/testing/async_sort.cu
@@ -15,6 +15,8 @@
 
 #  include <unittest/unittest.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 enum wait_policy
 {
   wait_for_futures,

diff --git a/thrust/testing/async_transform.cu b/thrust/testing/async_transform.cu
@@ -10,6 +10,8 @@
 #  include <unittest/unittest.h>
 #  include <unittest/util_async.h>
 
+_CCCL_SUPPRESS_DEPRECATED_PUSH
+
 template <typename T>
 struct divide_by_2
 {

diff --git a/thrust/thrust/async/copy.h b/thrust/thrust/async/copy.h
@@ -50,7 +50,7 @@ namespace unimplemented
 {
 
 template <typename FromPolicy, typename ToPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
-_CCCL_HOST event<FromPolicy> async_copy(
+CCCL_DEPRECATED _CCCL_HOST event<FromPolicy> async_copy(
   thrust::execution_policy<FromPolicy>& from_exec,
   thrust::execution_policy<ToPolicy>& to_exec,
   ForwardIt first,
@@ -111,11 +111,15 @@ struct copy_fn final
               THRUST_FWD(output)))
 
               template <typename... Args>
-              _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
+              CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
+    THRUST_RETURNS(call(THRUST_FWD(args)...))
 };
 
 } // namespace copy_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT copy_detail::copy_fn copy{};
 
 /*! \endcond

diff --git a/thrust/thrust/async/for_each.h b/thrust/thrust/async/for_each.h
@@ -50,7 +50,7 @@ namespace unimplemented
 {
 
 template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename UnaryFunction>
-_CCCL_HOST event<DerivedPolicy>
+CCCL_DEPRECATED _CCCL_HOST event<DerivedPolicy>
 async_for_each(thrust::execution_policy<DerivedPolicy>&, ForwardIt, Sentinel, UnaryFunction)
 {
   THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation<ForwardIt, false>::value),
@@ -88,11 +88,15 @@ struct for_each_fn final
           THRUST_FWD(f)))
 
           template <typename... Args>
-          _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
+          CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
+    THRUST_RETURNS(call(THRUST_FWD(args)...))
 };
 
 } // namespace for_each_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT for_each_detail::for_each_fn for_each{};
 
 /*! \endcond

diff --git a/thrust/thrust/async/reduce.h b/thrust/thrust/async/reduce.h
@@ -52,7 +52,7 @@ namespace unimplemented
 {
 
 template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename T, typename BinaryOp>
-_CCCL_HOST future<DerivedPolicy, T>
+CCCL_DEPRECATED _CCCL_HOST future<DerivedPolicy, T>
 async_reduce(thrust::execution_policy<DerivedPolicy>&, ForwardIt, Sentinel, T, BinaryOp)
 {
   THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation<ForwardIt, false>::value),
@@ -154,11 +154,15 @@ struct reduce_fn final
           thrust::plus<remove_cvref_t<typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type>>{}))
 
           template <typename... Args>
-          _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
+          CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
+    THRUST_RETURNS(call(THRUST_FWD(args)...))
 };
 
 } // namespace reduce_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT reduce_detail::reduce_fn reduce{};
 
 ///////////////////////////////////////////////////////////////////////////////
@@ -167,7 +171,7 @@ namespace unimplemented
 {
 
 template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt, typename T, typename BinaryOp>
-_CCCL_HOST event<DerivedPolicy>
+CCCL_DEPRECATED _CCCL_HOST event<DerivedPolicy>
 async_reduce_into(thrust::execution_policy<DerivedPolicy>&, ForwardIt, Sentinel, OutputIt, T, BinaryOp)
 {
   THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation<ForwardIt, false>::value),
@@ -290,11 +294,15 @@ struct reduce_into_fn final
         thrust::is_execution_policy<thrust::remove_cvref_t<T1>>{}))
 
         template <typename... Args>
-        _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
+        CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
+    THRUST_RETURNS(call(THRUST_FWD(args)...))
 };
 
 } // namespace reduce_into_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT reduce_into_detail::reduce_into_fn reduce_into{};
 
 /*! \endcond

diff --git a/thrust/thrust/async/scan.h b/thrust/thrust/async/scan.h
@@ -51,7 +51,7 @@ namespace unimplemented
 {
 
 template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt, typename BinaryOp>
-event<DerivedPolicy>
+CCCL_DEPRECATED event<DerivedPolicy>
 async_inclusive_scan(thrust::execution_policy<DerivedPolicy>&, ForwardIt, Sentinel, OutputIt, BinaryOp)
 {
   THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation<ForwardIt, false>::value),
@@ -65,7 +65,7 @@ template <typename DerivedPolicy,
           typename OutputIt,
           typename InitialValueType,
           typename BinaryOp>
-event<DerivedPolicy> async_exclusive_scan(
+CCCL_DEPRECATED event<DerivedPolicy> async_exclusive_scan(
   thrust::execution_policy<DerivedPolicy>&, ForwardIt, Sentinel, OutputIt, InitialValueType, BinaryOp)
 {
   THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation<ForwardIt, false>::value),
@@ -85,11 +85,12 @@ using thrust::async::unimplemented::async_inclusive_scan;
 struct inclusive_scan_fn final
 {
   template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt, typename BinaryOp>
-  auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                  ForwardIt&& first,
-                  Sentinel&& last,
-                  OutputIt&& out,
-                  BinaryOp&& op) const
+  CCCL_DEPRECATED auto operator()(
+    thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+    ForwardIt&& first,
+    Sentinel&& last,
+    OutputIt&& out,
+    BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -99,10 +100,11 @@ struct inclusive_scan_fn final
       THRUST_FWD(op)))
 
       template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
-      auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                      ForwardIt&& first,
-                      Sentinel&& last,
-                      OutputIt&& out) const
+      CCCL_DEPRECATED
+    auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+                    ForwardIt&& first,
+                    Sentinel&& last,
+                    OutputIt&& out) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -117,12 +119,13 @@ struct inclusive_scan_fn final
                 typename OutputIt,
                 typename InitialValueType,
                 typename BinaryOp>
-      auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                      ForwardIt&& first,
-                      Sentinel&& last,
-                      OutputIt&& out,
-                      InitialValueType&& init,
-                      BinaryOp&& op) const
+      CCCL_DEPRECATED
+    auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+                    ForwardIt&& first,
+                    Sentinel&& last,
+                    OutputIt&& out,
+                    InitialValueType&& init,
+                    BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -137,7 +140,7 @@ struct inclusive_scan_fn final
                 typename OutputIt,
                 typename BinaryOp,
                 typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, BinaryOp&& op) const
+      CCCL_DEPRECATED auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -148,7 +151,7 @@ struct inclusive_scan_fn final
       THRUST_FWD(op)))
 
       template <typename ForwardIt, typename Sentinel, typename OutputIt>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const
+      CCCL_DEPRECATED auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -164,7 +167,8 @@ struct inclusive_scan_fn final
                 typename InitialValueType,
                 typename BinaryOp,
                 typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const
+      CCCL_DEPRECATED
+    auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_inclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -178,6 +182,9 @@ struct inclusive_scan_fn final
 
 } // namespace inclusive_scan_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT inclusive_scan_detail::inclusive_scan_fn inclusive_scan{};
 
 namespace exclusive_scan_detail
@@ -195,12 +202,13 @@ struct exclusive_scan_fn final
             typename OutputIt,
             typename InitialValueType,
             typename BinaryOp>
-  auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                  ForwardIt&& first,
-                  Sentinel&& last,
-                  OutputIt&& out,
-                  InitialValueType&& init,
-                  BinaryOp&& op) const
+  CCCL_DEPRECATED auto operator()(
+    thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+    ForwardIt&& first,
+    Sentinel&& last,
+    OutputIt&& out,
+    InitialValueType&& init,
+    BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -211,11 +219,12 @@ struct exclusive_scan_fn final
       THRUST_FWD(op)))
 
       template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt, typename InitialValueType>
-      auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                      ForwardIt&& first,
-                      Sentinel&& last,
-                      OutputIt&& out,
-                      InitialValueType&& init) const
+      CCCL_DEPRECATED
+    auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+                    ForwardIt&& first,
+                    Sentinel&& last,
+                    OutputIt&& out,
+                    InitialValueType&& init) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -226,10 +235,11 @@ struct exclusive_scan_fn final
       thrust::plus<>{}))
 
       template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
-      auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
-                      ForwardIt&& first,
-                      Sentinel&& last,
-                      OutputIt&& out) const
+      CCCL_DEPRECATED
+    auto operator()(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
+                    ForwardIt&& first,
+                    Sentinel&& last,
+                    OutputIt&& out) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
@@ -245,7 +255,8 @@ struct exclusive_scan_fn final
                 typename InitialValueType,
                 typename BinaryOp,
                 typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const
+      CCCL_DEPRECATED
+    auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -261,7 +272,7 @@ struct exclusive_scan_fn final
                 typename OutputIt,
                 typename InitialValueType,
                 typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init) const
+      CCCL_DEPRECATED auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -273,7 +284,7 @@ struct exclusive_scan_fn final
       thrust::plus<>{}))
 
       template <typename ForwardIt, typename Sentinel, typename OutputIt>
-      auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const
+      CCCL_DEPRECATED auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const
     // ADL dispatch.
     THRUST_RETURNS(async_exclusive_scan(
       thrust::detail::select_system(iterator_system_t<remove_cvref_t<ForwardIt>>{},
@@ -287,6 +298,9 @@ struct exclusive_scan_fn final
 
 } // namespace exclusive_scan_detail
 
+// note: cannot add a CCCL_DEPRECATED here because the global variable is emitted into cudafe1.stub.c and we cannot
+// suppress the warning there
+//! deprecated [Since 2.8.0]
 _CCCL_GLOBAL_CONSTANT exclusive_scan_detail::exclusive_scan_fn exclusive_scan{};
 
 } // namespace async