Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Backport to 2.8: Deprecate thrust::async (#3324) #3388

Open
wants to merge 2 commits into
base: branch/2.8.x
Choose a base branch
from
+329 −228
Open

Backport to 2.8: Deprecate thrust::async (#3324) #3388

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
638a690
Deprecate thrust::async (#3324)
bernhardmgruber Jan 15, 2025
b61a97d
nvcc 11.1 fixes
bernhardmgruber Jan 15, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 2 additions & 0 deletions thrust/testing/async/exclusive_scan/mixin.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -93,12 +93,14 @@ struct simple
PostfixArgTuple&& postfix_tuple,
std::index_sequence<PostfixArgIndices...>)
{
_CCCL_SUPPRESS_DEPRECATED_PUSH
auto e = thrust::async::exclusive_scan(
std::get<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand Down
4 changes: 4 additions & 0 deletions thrust/testing/async/exclusive_scan/using_vs_adl.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -69,13 +69,15 @@ 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<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand All @@ -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<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand Down
2 changes: 2 additions & 0 deletions thrust/testing/async/inclusive_scan/mixin.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -109,12 +109,14 @@ struct simple
PostfixArgTuple&& postfix_tuple,
std::index_sequence<PostfixArgIndices...>)
{
_CCCL_SUPPRESS_DEPRECATED_PUSH
auto e = thrust::async::inclusive_scan(
std::get<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand Down
4 changes: 4 additions & 0 deletions thrust/testing/async/inclusive_scan/using_vs_adl.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -69,13 +69,15 @@ 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<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand All @@ -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<PrefixArgIndices>(THRUST_FWD(prefix_tuple))...,
input.cbegin(),
input.cend(),
output.begin(),
std::get<PostfixArgIndices>(THRUST_FWD(postfix_tuple))...);
_CCCL_SUPPRESS_DEPRECATED_POP
return e;
}
};
Expand Down
2 changes: 2 additions & 0 deletions thrust/testing/async_copy.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -10,6 +10,8 @@
# include <unittest/unittest.h>
# include <unittest/util_async.h>

_CCCL_SUPPRESS_DEPRECATED_PUSH

Comment on lines +13 to +14
Copy link
Collaborator

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

I am missing the _CCCL_SUPRESS_DEPRECATEED_POP

Copy link
Contributor Author

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

IIRC, I omitted it on purpose to suppress warnings in the cudafe1.c file. But I should have added a comment about it.

Can we still go ahead, given this is the backport PR?

# define DEFINE_ASYNC_COPY_CALLABLE(name, ...) \
struct THRUST_PP_CAT2(name, _fn) \
{ \
Expand Down
3 changes: 2 additions & 1 deletion thrust/testing/async_for_each.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -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) \
{ \
Expand All @@ -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
Expand Down
6 changes: 6 additions & 0 deletions thrust/testing/async_reduce.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -12,6 +12,8 @@
# include <unittest/unittest.h>
# include <unittest/util_async.h>

_CCCL_SUPPRESS_DEPRECATED_PUSH

template <typename T>
struct custom_plus
{
Expand Down Expand Up @@ -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.
Expand Down
2 changes: 2 additions & 0 deletions thrust/testing/async_reduce_into.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -13,6 +13,8 @@
# include <unittest/unittest.h>
# include <unittest/util_async.h>

_CCCL_SUPPRESS_DEPRECATED_PUSH

template <typename T>
struct custom_plus
{
Expand Down
2 changes: 2 additions & 0 deletions thrust/testing/async_sort.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -15,6 +15,8 @@

# include <unittest/unittest.h>

_CCCL_SUPPRESS_DEPRECATED_PUSH

enum wait_policy
{
wait_for_futures,
Expand Down
6 changes: 6 additions & 0 deletions thrust/testing/async_transform.cu
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -10,6 +10,8 @@
# include <unittest/unittest.h>
# include <unittest/util_async.h>

_CCCL_SUPPRESS_DEPRECATED_PUSH

template <typename T>
struct divide_by_2
{
Expand Down Expand Up @@ -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.
Expand Down
54 changes: 29 additions & 25 deletions thrust/thrust/async/copy.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -51,7 +51,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,
Expand All @@ -72,6 +72,7 @@ using thrust::async::unimplemented::async_copy;

struct copy_fn final
{
_CCCL_SUPPRESS_DEPRECATED_PUSH
template <typename FromPolicy, typename ToPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST static auto
call(thrust::detail::execution_policy_base<FromPolicy> const& from_exec,
Expand All @@ -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 <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST static auto call(thrust::detail::execution_policy_base<DerivedPolicy> 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<decltype(thrust::detail::derived_cast(thrust::detail::strip_const(exec)))>::tag_type{},
THRUST_FWD(output))) _CCCL_SUPPRESS_DEPRECATED_POP

template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename OutputIt>
_CCCL_HOST static auto call(thrust::detail::execution_policy_base<DerivedPolicy> 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<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_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_FWD(first),
THRUST_FWD(last),
THRUST_FWD(output)))

template <typename... Args>
_CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
template <typename... 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
Expand Down
30 changes: 17 additions & 13 deletions thrust/thrust/async/for_each.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -51,7 +51,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),
Expand All @@ -68,6 +68,7 @@ using thrust::async::unimplemented::async_for_each;

struct for_each_fn final
{
_CCCL_SUPPRESS_DEPRECATED_PUSH
template <typename DerivedPolicy, typename ForwardIt, typename Sentinel, typename UnaryFunction>
_CCCL_HOST static auto
call(thrust::detail::execution_policy_base<DerivedPolicy> const& exec,
Expand All @@ -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 <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_FWD(first),
THRUST_FWD(last),
THRUST_FWD(f)))

template <typename... Args>
_CCCL_NODISCARD _CCCL_HOST auto operator()(Args&&... args) const THRUST_RETURNS(call(THRUST_FWD(args)...))
THRUST_FWD(f))) _CCCL_SUPPRESS_DEPRECATED_POP

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_FWD(first),
THRUST_FWD(last),
THRUST_FWD(f)))

template <typename... 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
Expand Down
Loading
Loading