diff --git a/thrust/testing/async/exclusive_scan/mixin.h b/thrust/testing/async/exclusive_scan/mixin.h index bb6b4fe7aca..2ba787c6c8e 100644 --- a/thrust/testing/async/exclusive_scan/mixin.h +++ b/thrust/testing/async/exclusive_scan/mixin.h @@ -93,12 +93,14 @@ struct simple PostfixArgTuple&& postfix_tuple, std::index_sequence) { + _CCCL_SUPPRESS_DEPRECATED_PUSH auto e = thrust::async::exclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., input.cbegin(), input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; diff --git a/thrust/testing/async/exclusive_scan/using_vs_adl.cu b/thrust/testing/async/exclusive_scan/using_vs_adl.cu index 003136bc50e..5cfae148622 100644 --- a/thrust/testing/async/exclusive_scan/using_vs_adl.cu +++ b/thrust/testing/async/exclusive_scan/using_vs_adl.cu @@ -69,6 +69,7 @@ struct using_namespace // Importing the CPO into the current namespace should unambiguously resolve // this call to the CPO, as opposed to resolving to the thrust:: algorithm // via ADL. This is verified by checking that an event is returned. + _CCCL_SUPPRESS_DEPRECATED_PUSH using namespace thrust::async; thrust::device_event e = exclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., @@ -76,6 +77,7 @@ struct using_namespace input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; @@ -100,12 +102,14 @@ struct using_cpo // this call to the CPO, as opposed to resolving to the thrust:: algorithm // via ADL. This is verified by checking that an event is returned. using thrust::async::exclusive_scan; + _CCCL_SUPPRESS_DEPRECATED_PUSH thrust::device_event e = exclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., input.cbegin(), input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; diff --git a/thrust/testing/async/inclusive_scan/mixin.h b/thrust/testing/async/inclusive_scan/mixin.h index f3538f5eb94..453b5583681 100644 --- a/thrust/testing/async/inclusive_scan/mixin.h +++ b/thrust/testing/async/inclusive_scan/mixin.h @@ -109,12 +109,14 @@ struct simple PostfixArgTuple&& postfix_tuple, std::index_sequence) { + _CCCL_SUPPRESS_DEPRECATED_PUSH auto e = thrust::async::inclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., input.cbegin(), input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; diff --git a/thrust/testing/async/inclusive_scan/using_vs_adl.cu b/thrust/testing/async/inclusive_scan/using_vs_adl.cu index ff7270ad065..2713ecd3aaf 100644 --- a/thrust/testing/async/inclusive_scan/using_vs_adl.cu +++ b/thrust/testing/async/inclusive_scan/using_vs_adl.cu @@ -69,6 +69,7 @@ struct using_namespace // Importing the CPO into the current namespace should unambiguously resolve // this call to the CPO, as opposed to resolving to the thrust:: algorithm // via ADL. This is verified by checking that an event is returned. + _CCCL_SUPPRESS_DEPRECATED_PUSH using namespace thrust::async; thrust::device_event e = inclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., @@ -76,6 +77,7 @@ struct using_namespace input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; @@ -100,12 +102,14 @@ struct using_cpo // this call to the CPO, as opposed to resolving to the thrust:: algorithm // via ADL. This is verified by checking that an event is returned. using thrust::async::inclusive_scan; + _CCCL_SUPPRESS_DEPRECATED_PUSH thrust::device_event e = inclusive_scan( std::get(THRUST_FWD(prefix_tuple))..., input.cbegin(), input.cend(), output.begin(), std::get(THRUST_FWD(postfix_tuple))...); + _CCCL_SUPPRESS_DEPRECATED_POP return e; } }; diff --git a/thrust/testing/async_copy.cu b/thrust/testing/async_copy.cu index cf67652ac06..f3a4fd0ee41 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..2d471c55aca 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 { @@ -539,12 +541,16 @@ struct test_async_reduce_using // When you import the customization points into the global namespace, // they should be selected instead of the synchronous algorithms. { + _CCCL_SUPPRESS_DEPRECATED_PUSH using namespace thrust::async; f0a = reduce(d0a.begin(), d0a.end()); + _CCCL_SUPPRESS_DEPRECATED_POP } { + _CCCL_SUPPRESS_DEPRECATED_PUSH using thrust::async::reduce; f0b = reduce(d0b.begin(), d0b.end()); + _CCCL_SUPPRESS_DEPRECATED_POP } // ADL should find the synchronous algorithms. 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 0f94f9d88f3..03d332563d3 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 { @@ -401,12 +403,16 @@ struct test_async_transform_using // When you import the customization points into the global namespace, // they should be selected instead of the synchronous algorithms. { + _CCCL_SUPPRESS_DEPRECATED_PUSH using namespace thrust::async; f0a = transform(d0a.begin(), d0a.end(), d1a.begin(), op); + _CCCL_SUPPRESS_DEPRECATED_POP } { + _CCCL_SUPPRESS_DEPRECATED_PUSH using thrust::async::transform; f0b = transform(d0b.begin(), d0b.end(), d1b.begin(), op); + _CCCL_SUPPRESS_DEPRECATED_POP } // ADL should find the synchronous algorithms. diff --git a/thrust/thrust/async/copy.h b/thrust/thrust/async/copy.h index f981e1ac17d..deedbea72e6 100644 --- a/thrust/thrust/async/copy.h +++ b/thrust/thrust/async/copy.h @@ -51,7 +51,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, @@ -72,6 +72,7 @@ using thrust::async::unimplemented::async_copy; struct copy_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& from_exec, @@ -85,38 +86,41 @@ struct copy_fn final thrust::detail::derived_cast(thrust::detail::strip_const(to_exec)), THRUST_FWD(first), THRUST_FWD(last), - THRUST_FWD(output))) - - template - _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& output) - THRUST_RETURNS(copy_fn::call( - thrust::detail::derived_cast(thrust::detail::strip_const(exec)) - // Synthesize a suitable new execution policy, because we don't want to - // try and extract twice from the one we were passed. - , - typename remove_cvref_t::tag_type{}, + THRUST_FWD(output))) _CCCL_SUPPRESS_DEPRECATED_POP + + template + _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& exec, + ForwardIt&& first, + Sentinel&& last, + OutputIt&& output) + THRUST_RETURNS(copy_fn::call( + thrust::detail::derived_cast(thrust::detail::strip_const(exec)) + // Synthesize a suitable new execution policy, because we don't want to + // try and extract twice from the one we were passed. + , + typename remove_cvref_t::tag_type{}, + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(output))) + + template + _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output) THRUST_RETURNS(copy_fn::call( + thrust::detail::select_system(typename thrust::iterator_system>::type{}), + thrust::detail::select_system(typename thrust::iterator_system>::type{}), THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(output))) - template - _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output) - THRUST_RETURNS(copy_fn::call( - thrust::detail::select_system(typename thrust::iterator_system>::type{}), - thrust::detail::select_system(typename thrust::iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(output))) - - template - _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...)) + template + 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 656d43d4f15..c99f1050bac 100644 --- a/thrust/thrust/async/for_each.h +++ b/thrust/thrust/async/for_each.h @@ -51,7 +51,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), @@ -68,6 +68,7 @@ using thrust::async::unimplemented::async_for_each; struct for_each_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& exec, @@ -78,22 +79,25 @@ struct for_each_fn final THRUST_RETURNS(async_for_each(thrust::detail::derived_cast(thrust::detail::strip_const(exec)), THRUST_FWD(first), THRUST_FWD(last), - THRUST_FWD(f))) - - template - _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, UnaryFunction&& f) - THRUST_RETURNS(for_each_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(f))) - - template - _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...)) + THRUST_FWD(f))) _CCCL_SUPPRESS_DEPRECATED_POP + + template + _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, UnaryFunction&& f) THRUST_RETURNS(for_each_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(f))) + + template + 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 0a56a2ba51c..c9f8af40061 100644 --- a/thrust/thrust/async/reduce.h +++ b/thrust/thrust/async/reduce.h @@ -53,7 +53,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), @@ -70,6 +70,7 @@ using thrust::async::unimplemented::async_reduce; struct reduce_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& exec, @@ -83,28 +84,28 @@ struct reduce_fn final THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(init), - THRUST_FWD(op))) - - template - _CCCL_HOST static auto call4( - thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - T&& init, - thrust::true_type) + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH template + _CCCL_HOST static auto call4( + thrust::detail::execution_policy_base const& exec, + ForwardIt&& first, + Sentinel&& last, + T&& init, + thrust::true_type) // ADL dispatch. THRUST_RETURNS(async_reduce( thrust::detail::derived_cast(thrust::detail::strip_const(exec)), THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(init), - thrust::plus>{})) + thrust::plus>{})) _CCCL_SUPPRESS_DEPRECATED_POP - template - _CCCL_HOST static auto call3(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - thrust::true_type) + _CCCL_SUPPRESS_DEPRECATED_PUSH template + _CCCL_HOST static auto call3(thrust::detail::execution_policy_base const& exec, + ForwardIt&& first, + Sentinel&& last, + thrust::true_type) // ADL dispatch. THRUST_RETURNS(async_reduce( thrust::detail::derived_cast(thrust::detail::strip_const(exec)), @@ -112,24 +113,25 @@ struct reduce_fn final THRUST_FWD(last), typename iterator_traits>::value_type{}, thrust::plus>::value_type>>{})) - - template - _CCCL_HOST static auto call4(ForwardIt&& first, Sentinel&& last, T&& init, BinaryOp&& op, thrust::false_type) - THRUST_RETURNS(reduce_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(init), - THRUST_FWD(op))) - - template - _CCCL_HOST static auto call3(ForwardIt&& first, Sentinel&& last, T&& init, thrust::false_type) - THRUST_RETURNS(reduce_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(init), - thrust::plus>{})) + _CCCL_SUPPRESS_DEPRECATED_POP + + template + _CCCL_HOST static auto call4(ForwardIt&& first, Sentinel&& last, T&& init, BinaryOp&& op, thrust::false_type) + THRUST_RETURNS(reduce_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(init), + THRUST_FWD(op))) + + template + _CCCL_HOST static auto call3(ForwardIt&& first, Sentinel&& last, T&& init, thrust::false_type) + THRUST_RETURNS(reduce_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(init), + thrust::plus>{})) // MSVC WAR: MSVC gets angsty and eats all available RAM when we try to detect // if T1 is an execution_policy by using SFINAE. Switching to a static @@ -155,11 +157,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{}; /////////////////////////////////////////////////////////////////////////////// @@ -168,7 +174,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), @@ -185,6 +191,7 @@ using thrust::async::unimplemented::async_reduce_into; struct reduce_into_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template _CCCL_HOST static auto call(thrust::detail::execution_policy_base const& exec, @@ -200,16 +207,17 @@ struct reduce_into_fn final THRUST_FWD(last), THRUST_FWD(output), THRUST_FWD(init), - THRUST_FWD(op))) - - template - _CCCL_HOST static auto call5( - thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& output, - T&& init, - thrust::true_type) + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + _CCCL_HOST static auto call5( + thrust::detail::execution_policy_base const& exec, + ForwardIt&& first, + Sentinel&& last, + OutputIt&& output, + T&& init, + thrust::true_type) // ADL dispatch. THRUST_RETURNS(async_reduce_into( thrust::detail::derived_cast(thrust::detail::strip_const(exec)), @@ -217,15 +225,16 @@ struct reduce_into_fn final THRUST_FWD(last), THRUST_FWD(output), THRUST_FWD(init), - thrust::plus>{})) - - template - _CCCL_HOST static auto call4( - thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& output, - thrust::true_type) + thrust::plus>{})) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + _CCCL_HOST static auto call4( + thrust::detail::execution_policy_base const& exec, + ForwardIt&& first, + Sentinel&& last, + OutputIt&& output, + thrust::true_type) // ADL dispatch. THRUST_RETURNS(async_reduce_into( thrust::detail::derived_cast(thrust::detail::strip_const(exec)), @@ -234,41 +243,41 @@ struct reduce_into_fn final THRUST_FWD(output), typename iterator_traits>::value_type{}, thrust::plus>::value_type>>{})) - - template - _CCCL_HOST static auto call5( - ForwardIt&& first, Sentinel&& last, OutputIt&& output, T&& init, BinaryOp&& op, thrust::false_type) - THRUST_RETURNS(reduce_into_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}, - typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(output), - THRUST_FWD(init), - THRUST_FWD(op))) - - template - _CCCL_HOST static auto call4( - ForwardIt&& first, Sentinel&& last, OutputIt&& output, T&& init, thrust::false_type) - THRUST_RETURNS(reduce_into_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}, - typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(output), - THRUST_FWD(init), - thrust::plus>{})) - - template - _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output) - THRUST_RETURNS(reduce_into_fn::call( - thrust::detail::select_system(typename iterator_system>::type{}, - typename iterator_system>::type{}), - THRUST_FWD(first), - THRUST_FWD(last), - THRUST_FWD(output), - typename iterator_traits>::value_type{}, - thrust::plus>::value_type>>{})) + _CCCL_SUPPRESS_DEPRECATED_POP + + template + _CCCL_HOST static auto call5( + ForwardIt&& first, Sentinel&& last, OutputIt&& output, T&& init, BinaryOp&& op, thrust::false_type) + THRUST_RETURNS(reduce_into_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}, + typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(output), + THRUST_FWD(init), + THRUST_FWD(op))) + + template + _CCCL_HOST static auto call4(ForwardIt&& first, Sentinel&& last, OutputIt&& output, T&& init, thrust::false_type) + THRUST_RETURNS(reduce_into_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}, + typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(output), + THRUST_FWD(init), + thrust::plus>{})) + + template + _CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output) + THRUST_RETURNS(reduce_into_fn::call( + thrust::detail::select_system(typename iterator_system>::type{}, + typename iterator_system>::type{}), + THRUST_FWD(first), + THRUST_FWD(last), + THRUST_FWD(output), + typename iterator_traits>::value_type{}, + thrust::plus>::value_type>>{})) // MSVC WAR: MSVC gets angsty and eats all available RAM when we try to detect // if T1 is an execution_policy by using SFINAE. Switching to a static @@ -291,11 +300,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 5f287b64660..dd3913d49f1 100644 --- a/thrust/thrust/async/scan.h +++ b/thrust/thrust/async/scan.h @@ -52,7 +52,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), @@ -66,7 +66,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,45 +85,51 @@ using thrust::async::unimplemented::async_inclusive_scan; // Implementation of the thrust::async::inclusive_scan CPO. struct inclusive_scan_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH 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)), THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(out), - THRUST_FWD(op))) - - template - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out) const + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + 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)), THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(out), - thrust::plus<>{})) - - template - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - InitialValueType&& init, - BinaryOp&& op) const + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + 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)), @@ -131,14 +137,15 @@ struct inclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - THRUST_FWD(op))) - - template >>> - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, BinaryOp&& op) const + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template >>> + 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>{}, @@ -146,10 +153,10 @@ struct inclusive_scan_fn final THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(out), - THRUST_FWD(op))) + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP - template - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const + _CCCL_SUPPRESS_DEPRECATED_PUSH template + 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>{}, @@ -157,15 +164,17 @@ struct inclusive_scan_fn final THRUST_FWD(first), THRUST_FWD(last), THRUST_FWD(out), - thrust::plus<>{})) - - template >>> - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template >>> + 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>{}, @@ -174,11 +183,14 @@ struct inclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - THRUST_FWD(op))) + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP }; } // 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 @@ -190,18 +202,20 @@ using thrust::async::unimplemented::async_exclusive_scan; // Implementation of the thrust::async::exclusive_scan CPO. struct exclusive_scan_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template - 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)), @@ -209,14 +223,16 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - THRUST_FWD(op))) - - template - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out, - InitialValueType&& init) const + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + 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)), @@ -224,13 +240,15 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - thrust::plus<>{})) - - template - auto operator()(thrust::detail::execution_policy_base const& exec, - ForwardIt&& first, - Sentinel&& last, - OutputIt&& out) const + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template + 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)), @@ -238,15 +256,17 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), iterator_value_t>{}, - thrust::plus<>{})) - - template >>> - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init, BinaryOp&& op) const + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template >>> + 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>{}, @@ -255,14 +275,15 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - THRUST_FWD(op))) - - template >>> - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out, InitialValueType&& init) const + THRUST_FWD(op))) _CCCL_SUPPRESS_DEPRECATED_POP + + _CCCL_SUPPRESS_DEPRECATED_PUSH + template >>> + 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>{}, @@ -271,10 +292,10 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), THRUST_FWD(init), - thrust::plus<>{})) + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP - template - auto operator()(ForwardIt&& first, Sentinel&& last, OutputIt&& out) const + _CCCL_SUPPRESS_DEPRECATED_PUSH template + 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>{}, @@ -283,11 +304,14 @@ struct exclusive_scan_fn final THRUST_FWD(last), THRUST_FWD(out), iterator_value_t>{}, - thrust::plus<>{})) + thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP }; } // 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 7b76c0f9c00..42f07a501af 100644 --- a/thrust/thrust/async/sort.h +++ b/thrust/thrust/async/sort.h @@ -53,7 +53,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), @@ -71,11 +71,13 @@ using thrust::async::unimplemented::async_stable_sort; // clang-format off struct stable_sort_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template < typename DerivedPolicy , typename ForwardIt, typename Sentinel, typename StrictWeakOrdering > _CCCL_HOST + static auto call( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last @@ -89,12 +91,15 @@ struct stable_sort_fn final , THRUST_FWD(comp) ) ) + _CCCL_SUPPRESS_DEPRECATED_POP +_CCCL_SUPPRESS_DEPRECATED_PUSH template < typename DerivedPolicy , typename ForwardIt, typename Sentinel > _CCCL_HOST + static auto call( thrust::detail::execution_policy_base const& exec , ForwardIt&& first, Sentinel&& last @@ -109,6 +114,7 @@ struct stable_sort_fn final >{} ) ) + _CCCL_SUPPRESS_DEPRECATED_POP template _CCCL_HOST @@ -136,8 +142,11 @@ struct stable_sort_fn final ) template + #if !_CCCL_CUDA_COMPILER(CLANG) + // clang in CUDA mode can only handle one attribute _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + #endif + CCCL_DEPRECATED auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -146,13 +155,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)); @@ -168,6 +180,7 @@ using thrust::async::fallback::async_sort; // clang-format off struct sort_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template < typename DerivedPolicy , typename ForwardIt, typename Sentinel, typename StrictWeakOrdering @@ -186,6 +199,7 @@ struct sort_fn final , THRUST_FWD(comp) ) ) + _CCCL_SUPPRESS_DEPRECATED_POP template < typename DerivedPolicy @@ -249,8 +263,11 @@ struct sort_fn final ) template + #if !_CCCL_CUDA_COMPILER(CLANG) + // clang in CUDA mode can only handle one attribute _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + #endif + CCCL_DEPRECATED auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -259,6 +276,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 0e97e0bce51..cb2a4a05217 100644 --- a/thrust/thrust/async/transform.h +++ b/thrust/thrust/async/transform.h @@ -51,7 +51,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), @@ -69,6 +69,7 @@ using thrust::async::unimplemented::async_transform; // clang-format off struct transform_fn final { + _CCCL_SUPPRESS_DEPRECATED_PUSH template < typename DerivedPolicy , typename ForwardIt, typename Sentinel, typename OutputIt @@ -91,6 +92,7 @@ struct transform_fn final , THRUST_FWD(op) ) ) + _CCCL_SUPPRESS_DEPRECATED_POP template < typename ForwardIt, typename Sentinel, typename OutputIt @@ -115,8 +117,8 @@ struct transform_fn final ) template - _CCCL_NODISCARD _CCCL_HOST - auto operator()(Args&&... args) const + CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST + auto operator()(Args&&... args) const THRUST_RETURNS( call(THRUST_FWD(args)...) ) @@ -125,6 +127,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