From 80bef1d4b79a64b5e4473221e2f49e4161e4e56e Mon Sep 17 00:00:00 2001 From: Bernhard Manfred Gruber Date: Fri, 10 Jan 2025 15:08:14 +0100 Subject: [PATCH] Deprecate thrust::async Fixes: #100 --- thrust/testing/async_copy.cu | 2 + thrust/testing/async_for_each.cu | 3 +- thrust/testing/async_reduce.cu | 2 + thrust/testing/async_reduce_into.cu | 2 + thrust/testing/async_sort.cu | 2 + thrust/testing/async_transform.cu | 2 + thrust/thrust/async/copy.h | 8 ++- thrust/thrust/async/for_each.h | 8 ++- thrust/thrust/async/reduce.h | 16 +++-- thrust/thrust/async/scan.h | 90 +++++++++++++++++------------ thrust/thrust/async/sort.h | 14 +++-- thrust/thrust/async/transform.h | 7 ++- 12 files changed, 103 insertions(+), 53 deletions(-) diff --git a/thrust/testing/async_copy.cu b/thrust/testing/async_copy.cu index cd18d83a782..ce3451dc829 100644 --- a/thrust/testing/async_copy.cu +++ b/thrust/testing/async_copy.cu @@ -10,6 +10,8 @@ # include # include +_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 index a4ca2771b72..f939fe85c7e 100644 --- a/thrust/testing/async_for_each.cu +++ b/thrust/testing/async_for_each.cu @@ -8,6 +8,8 @@ # include +_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 index 85fcb3ef395..8ee48856c97 100644 --- a/thrust/testing/async_reduce.cu +++ b/thrust/testing/async_reduce.cu @@ -12,6 +12,8 @@ # include # include +_CCCL_SUPPRESS_DEPRECATED_PUSH + template struct custom_plus { diff --git a/thrust/testing/async_reduce_into.cu b/thrust/testing/async_reduce_into.cu index e757ac3a708..3cde0e68f6d 100644 --- a/thrust/testing/async_reduce_into.cu +++ b/thrust/testing/async_reduce_into.cu @@ -13,6 +13,8 @@ # include # include +_CCCL_SUPPRESS_DEPRECATED_PUSH + template struct custom_plus { diff --git a/thrust/testing/async_sort.cu b/thrust/testing/async_sort.cu index 77144779814..b7de68fceca 100644 --- a/thrust/testing/async_sort.cu +++ b/thrust/testing/async_sort.cu @@ -15,6 +15,8 @@ # include +_CCCL_SUPPRESS_DEPRECATED_PUSH + enum wait_policy { wait_for_futures, diff --git a/thrust/testing/async_transform.cu b/thrust/testing/async_transform.cu index bfb30006ff2..4f7c02bba4c 100644 --- a/thrust/testing/async_transform.cu +++ b/thrust/testing/async_transform.cu @@ -10,6 +10,8 @@ # include # include +_CCCL_SUPPRESS_DEPRECATED_PUSH + template struct divide_by_2 { diff --git a/thrust/thrust/async/copy.h b/thrust/thrust/async/copy.h index 1adc90c3dff..1e003fdecd2 100644 --- a/thrust/thrust/async/copy.h +++ b/thrust/thrust/async/copy.h @@ -50,7 +50,7 @@ namespace unimplemented { template -_CCCL_HOST event async_copy( +CCCL_DEPRECATED _CCCL_HOST event async_copy( thrust::execution_policy& from_exec, thrust::execution_policy& to_exec, ForwardIt first, @@ -111,11 +111,15 @@ struct copy_fn final THRUST_FWD(output))) template - _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 index 6128b6a7625..8f064636bf9 100644 --- a/thrust/thrust/async/for_each.h +++ b/thrust/thrust/async/for_each.h @@ -50,7 +50,7 @@ namespace unimplemented { template -_CCCL_HOST event +CCCL_DEPRECATED _CCCL_HOST event async_for_each(thrust::execution_policy&, ForwardIt, Sentinel, UnaryFunction) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -88,11 +88,15 @@ struct for_each_fn final THRUST_FWD(f))) template - _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 index a6cea12b5ab..f958efd0fb7 100644 --- a/thrust/thrust/async/reduce.h +++ b/thrust/thrust/async/reduce.h @@ -52,7 +52,7 @@ namespace unimplemented { template -_CCCL_HOST future +CCCL_DEPRECATED _CCCL_HOST future async_reduce(thrust::execution_policy&, ForwardIt, Sentinel, T, BinaryOp) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -154,11 +154,15 @@ struct reduce_fn final thrust::plus>::value_type>>{})) template - _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 -_CCCL_HOST event +CCCL_DEPRECATED _CCCL_HOST event async_reduce_into(thrust::execution_policy&, ForwardIt, Sentinel, OutputIt, T, BinaryOp) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -290,11 +294,15 @@ struct reduce_into_fn final thrust::is_execution_policy>{})) template - _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 index 4963d17225e..3625b12c76f 100644 --- a/thrust/thrust/async/scan.h +++ b/thrust/thrust/async/scan.h @@ -51,7 +51,7 @@ namespace unimplemented { template -event +CCCL_DEPRECATED event async_inclusive_scan(thrust::execution_policy&, ForwardIt, Sentinel, OutputIt, BinaryOp) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -65,7 +65,7 @@ template -event async_exclusive_scan( +CCCL_DEPRECATED event async_exclusive_scan( thrust::execution_policy&, ForwardIt, Sentinel, OutputIt, InitialValueType, BinaryOp) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -85,11 +85,12 @@ using thrust::async::unimplemented::async_inclusive_scan; struct inclusive_scan_fn final { template - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - BinaryOp&& op) const + CCCL_DEPRECATED auto operator()( + thrust::detail::execution_policy_base 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 - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out) const + CCCL_DEPRECATED + auto operator()(thrust::detail::execution_policy_base 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 const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - InitialValueType&& init, - BinaryOp&& op) const + CCCL_DEPRECATED + auto operator()(thrust::detail::execution_policy_base 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>>> - 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>{}, @@ -148,7 +151,7 @@ struct inclusive_scan_fn final THRUST_FWD(op))) template - 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>{}, @@ -164,7 +167,8 @@ struct inclusive_scan_fn final typename InitialValueType, typename BinaryOp, typename = std::enable_if_t>>> - 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>{}, @@ -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 const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - InitialValueType&& init, - BinaryOp&& op) const + CCCL_DEPRECATED auto operator()( + thrust::detail::execution_policy_base 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 - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - InitialValueType&& init) const + CCCL_DEPRECATED + auto operator()(thrust::detail::execution_policy_base 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 - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out) const + CCCL_DEPRECATED + auto operator()(thrust::detail::execution_policy_base 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>>> - 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>{}, @@ -261,7 +272,7 @@ struct exclusive_scan_fn final typename OutputIt, typename InitialValueType, typename = std::enable_if_t>>> - 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>{}, @@ -273,7 +284,7 @@ struct exclusive_scan_fn final thrust::plus<>{})) template - 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>{}, @@ -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 diff --git a/thrust/thrust/async/sort.h b/thrust/thrust/async/sort.h index ae37abb5d61..a3adfa504cc 100644 --- a/thrust/thrust/async/sort.h +++ b/thrust/thrust/async/sort.h @@ -52,7 +52,7 @@ namespace unimplemented { template -_CCCL_HOST event +CCCL_DEPRECATED _CCCL_HOST event async_stable_sort(thrust::execution_policy&, ForwardIt, Sentinel, StrictWeakOrdering) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -136,7 +136,7 @@ struct stable_sort_fn final template _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + CCCL_DEPRECATED auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -145,13 +145,16 @@ struct stable_sort_fn final } // namespace stable_sort_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 stable_sort_detail::stable_sort_fn stable_sort{}; namespace fallback { template -_CCCL_HOST event +CCCL_DEPRECATED _CCCL_HOST event async_sort(thrust::execution_policy& exec, ForwardIt&& first, Sentinel&& last, StrictWeakOrdering&& comp) { return async_stable_sort(thrust::detail::derived_cast(exec), THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(comp)); @@ -249,7 +252,7 @@ struct sort_fn final template _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + CCCL_DEPRECATED auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -258,6 +261,9 @@ struct sort_fn final } // namespace sort_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 sort_detail::sort_fn sort{}; /*! \endcond diff --git a/thrust/thrust/async/transform.h b/thrust/thrust/async/transform.h index 0862141ee0d..f5ed6787ed9 100644 --- a/thrust/thrust/async/transform.h +++ b/thrust/thrust/async/transform.h @@ -50,7 +50,7 @@ namespace unimplemented { template -_CCCL_HOST event async_transform( +CCCL_DEPRECATED _CCCL_HOST event async_transform( thrust::execution_policy& exec, ForwardIt first, Sentinel last, OutputIt output, UnaryOperation op) { THRUST_STATIC_ASSERT_MSG((thrust::detail::depend_on_instantiation::value), @@ -115,7 +115,7 @@ struct transform_fn final template _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + CCCL_DEPRECATED auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -124,6 +124,9 @@ struct transform_fn final } // namespace transform_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 transform_detail::transform_fn transform{}; /*! \endcond