-
Notifications
You must be signed in to change notification settings - Fork 106
Update cuco implementations to use cuda::std utilities when appropriate #708
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.
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
Conversation
bdice
left a comment
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Generally LGTM -- ideally we would use cuda::std::tuple in the places where thrust::tuple is being used.
| #include <cuco/detail/bitwise_compare.cuh> | ||
| #include <cuco/detail/pair/traits.hpp> | ||
|
|
||
| #include <thrust/tuple.h> |
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
I think cuda::std::tuple should be preferred here? Please try it. It's also okay to defer this to another PR.
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Right, the issue is from the outdated CCCL version, where zip_iterator is incompatible with cuda::std::tuple: https://godbolt.org/z/jnKafGPf4. Currently, cuco is still pulling in rapids-cmake branch-25.02, which brings in CCCL 2.7. Once we update to a newer version of rapids-cmake, CCCL will be upgraded accordingly, and we’ll be able to fully replace thrust::tuple with cuda::std::tuple in cuco.
This PR updates cuco implementations to use
cuda::stdutilities in place ofstdin device code and to replacethrustutilities wherever possible, as they are being deprecated in CCCL.