Skip to content

Commit

Permalink
Redefine and deprecate thrust::remove_cvref (#3394)
Browse files Browse the repository at this point in the history 
* Redefine and deprecate thrust::remove_cvref

Co-authored-by: Michael Schellenberger Costa <[email protected]>
  • Loading branch information
@bernhardmgruber @miscco
bernhardmgruber and miscco authored Jan 16, 2025
1 parent 3267f42 commit 9abf805
Show file tree
Hide file tree
Showing 23 changed files with 130 additions and 165 deletions.
2 changes: 1 addition & 1 deletion thrust/testing/async/exclusive_scan/using_vs_adl.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -29,7 +29,7 @@ struct adl_host_synchronous
thrust::host_vector<input_value_type> host_input(input.cbegin(), input.cend());
thrust::host_vector<output_value_type> host_output(host_input.size());

using OutIter = thrust::remove_cvref_t<decltype(host_output.begin())>;
using OutIter = cuda::std::remove_cvref_t<decltype(host_output.begin())>;

// ADL should resolve this to the synchronous `thrust::` algorithm.
// This is checked by ensuring that the call returns an output iterator.
Expand Down
2 changes: 1 addition & 1 deletion thrust/testing/async/inclusive_scan/using_vs_adl.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -29,7 +29,7 @@ struct adl_host_synchronous
thrust::host_vector<input_value_type> host_input(input.cbegin(), input.cend());
thrust::host_vector<output_value_type> host_output(host_input.size());

using OutIter = thrust::remove_cvref_t<decltype(host_output.begin())>;
using OutIter = cuda::std::remove_cvref_t<decltype(host_output.begin())>;

// ADL should resolve this to the synchronous `thrust::` algorithm.
// This is checked by ensuring that the call returns an output iterator.
Expand Down
2 changes: 1 addition & 1 deletion thrust/testing/async/test_policy_overloads.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -147,7 +147,7 @@ struct test_policy_overloads
{
// Sink the prefix tuple into a const local so it can be safely passed to
// multiple invocations without worrying about potential modifications.
using prefix_tuple_type = thrust::remove_cvref_t<PrefixArgTuple>;
using prefix_tuple_type = cuda::std::remove_cvref_t<PrefixArgTuple>;
prefix_tuple_type const prefix_tuple = THRUST_FWD(prefix_tuple_ref);

using postfix_tuple_type = std::tuple_element_t<PostfixIdx, postfix_args_type>;
Expand Down
11 changes: 7 additions & 4 deletions thrust/thrust/async/copy.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -36,7 +36,8 @@
# include <thrust/detail/static_assert.h>
# include <thrust/event.h>
# include <thrust/system/detail/adl/async/copy.h>
# include <thrust/type_traits/remove_cvref.h>

# include <cuda/std/type_traits>

THRUST_NAMESPACE_BEGIN

Expand Down Expand Up @@ -96,15 +97,17 @@ struct copy_fn final
// 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<decltype(thrust::detail::derived_cast(thrust::detail::strip_const(exec)))>::tag_type{},
typename ::cuda::std::remove_cvref_t<
decltype(thrust::detail::derived_cast(thrust::detail::strip_const(exec)))>::tag_type{},
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output)))

template <typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output) THRUST_RETURNS(copy_fn::call(
thrust::detail::select_system(typename thrust::iterator_system<remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename thrust::iterator_system<remove_cvref_t<OutputIt>>::type{}),
thrust::detail::select_system(
typename thrust::iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename thrust::iterator_system<::cuda::std::remove_cvref_t<OutputIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output)))
Expand Down
5 changes: 3 additions & 2 deletions thrust/thrust/async/for_each.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -36,7 +36,8 @@
# include <thrust/detail/static_assert.h>
# include <thrust/event.h>
# include <thrust/system/detail/adl/async/for_each.h>
# include <thrust/type_traits/remove_cvref.h>

# include <cuda/std/type_traits>

THRUST_NAMESPACE_BEGIN

Expand Down Expand Up @@ -81,7 +82,7 @@ struct for_each_fn final

template <typename ForwardIt, typename Sentinel, typename UnaryFunction>
_CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, UnaryFunction&& f) THRUST_RETURNS(for_each_fn::call(
thrust::detail::select_system(typename iterator_system<remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(f)))
Expand Down
57 changes: 31 additions & 26 deletions thrust/thrust/async/reduce.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -38,7 +38,8 @@
# include <thrust/system/detail/adl/async/reduce.h>
# include <thrust/type_traits/is_execution_policy.h>
# include <thrust/type_traits/logical_metafunctions.h>
# include <thrust/type_traits/remove_cvref.h>

# include <cuda/std/type_traits>

THRUST_NAMESPACE_BEGIN

Expand Down Expand Up @@ -97,7 +98,7 @@ struct reduce_fn final
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(init),
thrust::plus<remove_cvref_t<T>>{})) _CCCL_SUPPRESS_DEPRECATED_POP
thrust::plus<::cuda::std::remove_cvref_t<T>>{})) _CCCL_SUPPRESS_DEPRECATED_POP

template <typename DerivedPolicy, typename ForwardIt, typename Sentinel>
_CCCL_HOST _CCCL_SUPPRESS_DEPRECATED_PUSH
Expand All @@ -110,14 +111,15 @@ struct reduce_fn final
thrust::detail::derived_cast(thrust::detail::strip_const(exec)),
THRUST_FWD(first),
THRUST_FWD(last),
typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<remove_cvref_t<typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type>>{}))
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<
::cuda::std::remove_cvref_t<typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type>>{}))
_CCCL_SUPPRESS_DEPRECATED_POP

template <typename ForwardIt, typename Sentinel, typename T, typename BinaryOp>
_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<remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(init),
Expand All @@ -126,34 +128,35 @@ struct reduce_fn final
template <typename ForwardIt, typename Sentinel, typename T>
_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<remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(init),
thrust::plus<remove_cvref_t<T>>{}))
thrust::plus<::cuda::std::remove_cvref_t<T>>{}))

// 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
// dispatch pattern to prevent this.
template <typename T1, typename T2, typename T3>
_CCCL_HOST static auto call(T1&& t1, T2&& t2, T3&& t3) THRUST_RETURNS(reduce_fn::call3(
THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), thrust::is_execution_policy<thrust::remove_cvref_t<T1>>{}))
THRUST_FWD(t1), THRUST_FWD(t2), THRUST_FWD(t3), thrust::is_execution_policy<::cuda::std::remove_cvref_t<T1>>{}))

template <typename T1, typename T2, typename T3, typename T4>
_CCCL_HOST static auto call(T1&& t1, T2&& t2, T3&& t3, T4&& t4) THRUST_RETURNS(reduce_fn::call4(
THRUST_FWD(t1),
THRUST_FWD(t2),
THRUST_FWD(t3),
THRUST_FWD(t4),
thrust::is_execution_policy<thrust::remove_cvref_t<T1>>{}))
thrust::is_execution_policy<::cuda::std::remove_cvref_t<T1>>{}))

template <typename ForwardIt, typename Sentinel>
_CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last) THRUST_RETURNS(reduce_fn::call(
thrust::detail::select_system(typename iterator_system<remove_cvref_t<ForwardIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<remove_cvref_t<typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type>>{}))
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<::cuda::std::remove_cvref_t<
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type>>{}))

template <typename... Args>
CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
Expand Down Expand Up @@ -222,7 +225,7 @@ struct reduce_into_fn final
THRUST_FWD(last),
THRUST_FWD(output),
THRUST_FWD(init),
thrust::plus<remove_cvref_t<T>>{})) _CCCL_SUPPRESS_DEPRECATED_POP
thrust::plus<::cuda::std::remove_cvref_t<T>>{})) _CCCL_SUPPRESS_DEPRECATED_POP

template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST _CCCL_SUPPRESS_DEPRECATED_PUSH
Expand All @@ -237,16 +240,17 @@ struct reduce_into_fn final
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output),
typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<remove_cvref_t<typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type>>{}))
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<
::cuda::std::remove_cvref_t<typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type>>{}))
_CCCL_SUPPRESS_DEPRECATED_POP

template <typename ForwardIt, typename Sentinel, typename OutputIt, typename T, typename BinaryOp>
_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<remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<remove_cvref_t<OutputIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<::cuda::std::remove_cvref_t<OutputIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output),
Expand All @@ -256,24 +260,25 @@ struct reduce_into_fn final
template <typename ForwardIt, typename Sentinel, typename OutputIt, typename T>
_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<remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<remove_cvref_t<OutputIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<::cuda::std::remove_cvref_t<OutputIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output),
THRUST_FWD(init),
thrust::plus<remove_cvref_t<T>>{}))
thrust::plus<::cuda::std::remove_cvref_t<T>>{}))

template <typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST static auto call(ForwardIt&& first, Sentinel&& last, OutputIt&& output)
THRUST_RETURNS(reduce_into_fn::call(
thrust::detail::select_system(typename iterator_system<remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<remove_cvref_t<OutputIt>>::type{}),
thrust::detail::select_system(typename iterator_system<::cuda::std::remove_cvref_t<ForwardIt>>::type{},
typename iterator_system<::cuda::std::remove_cvref_t<OutputIt>>::type{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output),
typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<remove_cvref_t<typename iterator_traits<remove_cvref_t<ForwardIt>>::value_type>>{}))
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::value_type{},
thrust::plus<::cuda::std::remove_cvref_t<
typename iterator_traits<::cuda::std::remove_cvref_t<ForwardIt>>::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
Expand All @@ -284,7 +289,7 @@ struct reduce_into_fn final
THRUST_FWD(t2),
THRUST_FWD(t3),
THRUST_FWD(t4),
thrust::is_execution_policy<thrust::remove_cvref_t<T1>>{}))
thrust::is_execution_policy<::cuda::std::remove_cvref_t<T1>>{}))

template <typename T1, typename T2, typename T3, typename T4, typename T5>
_CCCL_HOST static auto call(T1&& t1, T2&& t2, T3&& t3, T4&& t4, T5&& t5) THRUST_RETURNS(reduce_into_fn::call5(
Expand All @@ -293,7 +298,7 @@ struct reduce_into_fn final
THRUST_FWD(t3),
THRUST_FWD(t4),
THRUST_FWD(t5),
thrust::is_execution_policy<thrust::remove_cvref_t<T1>>{}))
thrust::is_execution_policy<::cuda::std::remove_cvref_t<T1>>{}))

template <typename... Args>
CCCL_DEPRECATED _CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const
Expand Down
39 changes: 20 additions & 19 deletions thrust/thrust/async/scan.h
View file
Original file line number Diff line number Diff line change
Expand Up @@ -39,7 +39,8 @@
# include <thrust/system/detail/adl/async/scan.h>
# include <thrust/type_traits/is_execution_policy.h>
# include <thrust/type_traits/logical_metafunctions.h>
# include <thrust/type_traits/remove_cvref.h>

# include <cuda/std/type_traits>

THRUST_NAMESPACE_BEGIN

Expand Down Expand Up @@ -139,13 +140,13 @@ struct inclusive_scan_fn final
typename Sentinel,
typename OutputIt,
typename BinaryOp,
typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
typename = std::enable_if_t<!is_execution_policy_v<::cuda::std::remove_cvref_t<ForwardIt>>>>
CCCL_DEPRECATED _CCCL_SUPPRESS_DEPRECATED_PUSH
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
Expand All @@ -156,8 +157,8 @@ struct inclusive_scan_fn final
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
Expand All @@ -168,13 +169,13 @@ struct inclusive_scan_fn final
typename OutputIt,
typename InitialValueType,
typename BinaryOp,
typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
typename = std::enable_if_t<!is_execution_policy_v<::cuda::std::remove_cvref_t<ForwardIt>>>>
CCCL_DEPRECATED _CCCL_SUPPRESS_DEPRECATED_PUSH
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
Expand Down Expand Up @@ -248,21 +249,21 @@ struct exclusive_scan_fn final
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
iterator_value_t<remove_cvref_t<ForwardIt>>{},
iterator_value_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP

template <typename ForwardIt,
typename Sentinel,
typename OutputIt,
typename InitialValueType,
typename BinaryOp,
typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
typename = std::enable_if_t<!is_execution_policy_v<::cuda::std::remove_cvref_t<ForwardIt>>>>
CCCL_DEPRECATED _CCCL_SUPPRESS_DEPRECATED_PUSH
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
Expand All @@ -273,13 +274,13 @@ struct exclusive_scan_fn final
typename Sentinel,
typename OutputIt,
typename InitialValueType,
typename = std::enable_if_t<!is_execution_policy_v<remove_cvref_t<ForwardIt>>>>
typename = std::enable_if_t<!is_execution_policy_v<::cuda::std::remove_cvref_t<ForwardIt>>>>
CCCL_DEPRECATED _CCCL_SUPPRESS_DEPRECATED_PUSH
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
Expand All @@ -291,12 +292,12 @@ struct exclusive_scan_fn final
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>>{},
iterator_system_t<remove_cvref_t<OutputIt>>{}),
thrust::detail::select_system(iterator_system_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
iterator_system_t<::cuda::std::remove_cvref_t<OutputIt>>{}),
THRUST_FWD(first),
THRUST_FWD(last),
THRUST_FWD(out),
iterator_value_t<remove_cvref_t<ForwardIt>>{},
iterator_value_t<::cuda::std::remove_cvref_t<ForwardIt>>{},
thrust::plus<>{})) _CCCL_SUPPRESS_DEPRECATED_POP
};

Expand Down
Loading

0 comments on commit 9abf805

Please sign in to comment.