Skip to content

Fix the new interface for swap in cuda thrust operators. - #39

Closed
hzhangxyz wants to merge 1 commit into
mainfrom
fix/cuda-12-9-error-about-swap-in-thrust
Closed

Fix the new interface for swap in cuda thrust operators.#39
hzhangxyz wants to merge 1 commit into
mainfrom
fix/cuda-12-9-error-about-swap-in-thrust

Conversation

@hzhangxyz

@hzhangxyz hzhangxyz commented Jun 16, 2025

Copy link
Copy Markdown
Member

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。

Copilot AI review requested due to automatic review settings June 16, 2025 13:36

This comment was marked as outdated.

@hzhangxyz
hzhangxyz force-pushed the fix/cuda-12-9-error-about-swap-in-thrust branch 3 times, most recently from 4b316ec to 9c190f6 Compare June 16, 2025 14:06
@hzhangxyz
hzhangxyz requested a review from Copilot June 16, 2025 14:18
@hzhangxyz

Copy link
Copy Markdown
Member Author

Wait, it seems this does not work.

@hzhangxyz
hzhangxyz marked this pull request as draft June 16, 2025 14:18

This comment was marked as outdated.

@hzhangxyz

Copy link
Copy Markdown
Member Author

I asked it there.

@hzhangxyz

hzhangxyz commented Jun 16, 2025

Copy link
Copy Markdown
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.

@hzhangxyz

hzhangxyz commented Jun 16, 2025

Copy link
Copy Markdown
Member Author

It seems we need to use cuda::std::array instead of std::array.

See: NVIDIA/cccl#5001 (comment)

@hzhangxyz
hzhangxyz force-pushed the fix/cuda-12-9-error-about-swap-in-thrust branch from 9c190f6 to 2016b1f Compare June 16, 2025 15:34
@hzhangxyz
hzhangxyz marked this pull request as ready for review June 16, 2025 15:35
@hzhangxyz
hzhangxyz requested a review from Copilot June 16, 2025 15:35

Copilot AI left a comment

Copy link
Copy Markdown

Choose a reason for hiding this comment

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

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
hzhangxyz requested review from BrevityD, CuSO4Deposit and stevapple and removed request for BrevityD and stevapple June 16, 2025 15:39
@hzhangxyz

Copy link
Copy Markdown
Member Author

Implementing cuda::std::swap does not solve this issue, so close it.

@hzhangxyz hzhangxyz closed this Jun 17, 2025
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

None yet

Development

Successfully merging this pull request may close these issues.

🐛 [BUG] - CUDA Compilation Errors after Upgrading to cuda 12.9

2 participants