Fix the new interface for swap in cuda thrust operators. - #39
Closed
hzhangxyz wants to merge 1 commit into
Closed
Conversation
hzhangxyz
force-pushed
the
fix/cuda-12-9-error-about-swap-in-thrust
branch
3 times, most recently
from
June 16, 2025 14:06
4b316ec to
9c190f6
Compare
Member
Author
|
Wait, it seems this does not work. |
hzhangxyz
marked this pull request as draft
June 16, 2025 14:18
Member
Author
|
I asked it there. |
Member
Author
|
I tried another solution by replace std::array to plain array, this is ugly. Let's see if is there any other solution these days. See branch: fix/cuda-12-9-error-about-thrust-sort-by-use-plain-array Note: This branch does not work currently. |
Member
Author
|
It seems we need to use |
hzhangxyz
force-pushed
the
fix/cuda-12-9-error-about-swap-in-thrust
branch
from
June 16, 2025 15:34
9c190f6 to
2016b1f
Compare
hzhangxyz
marked this pull request as ready for review
June 16, 2025 15:35
There was a problem hiding this comment.
Pull Request Overview
This PR updates the array types in the CUDA Hamiltonian interface to use cuda::std::array instead of std::array so that the custom swap logic can be used with the new thrust requirements.
- Replaced array type references in helper structs and device functions.
- Updated reinterpret_cast calls in thrust sort functions to use cuda::std::array pointers.
Comments suppressed due to low confidence (2)
qmb/_hamiltonian_cuda.cu:245
- Verify that the layout of cuda::std::array is compatible with reinterpret_cast operations used in thrust sorting. If necessary, ensure a helper function or static_assert is in place to validate compatibility.
reinterpret_cast<cuda::std::array<std::uint8_t, n_qubytes>*>(sorted_result_configs.data_ptr())
qmb/_hamiltonian_cuda.cu:622
- Confirm that reinterpret_cast casts to cuda::std::array pointers remain safe and valid when used in thrust::sort calls, ensuring consistency with the new cuda interface.
reinterpret_cast<cuda::std::array<std::uint8_t, n_qubytes>*>(sorted_exclude_configs.data_ptr())
hzhangxyz
requested review from
BrevityD,
CuSO4Deposit and
stevapple
and removed request for
BrevityD and
stevapple
June 16, 2025 15:39
Member
Author
|
Implementing cuda::std::swap does not solve this issue, so close it. |
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.This suggestion is invalid because no changes were made to the code.Suggestions cannot be applied while the pull request is closed.Suggestions cannot be applied while viewing a subset of changes.Only one suggestion per line can be applied in a batch.Add this suggestion to a batch that can be applied as a single commit.Applying suggestions on deleted lines is not supported.You must change the existing code in this line in order to create a valid suggestion.Outdated suggestions cannot be applied.This suggestion has been applied or marked resolved.Suggestions cannot be applied from pending reviews.Suggestions cannot be applied on multi-line comments.Suggestions cannot be applied while the pull request is queued to merge.Suggestion cannot be applied right now. Please check back later.
Description
The new thrust interface requires std::swap or cuda::std::swap for thrust::sort, instead of the previous plain assignments, so we need to implement cuda::std::swap manually since cuda 12.9.
Closes: #36
See: NVIDIA/cccl#3333
Checklist:
给Reviewer看的简单介绍:
cuda 12.9 开始, thrust所在CCCL更新至3.0,有很多API break了,其中影响到我们的是,thrust的sort函数调用了swap函数,而std::array之前通过内部手动交换没有问题,但是他的swap并没有GPU实现,所以nvcc叫了。
解决方法:用cuda::std::array。