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.

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

Replace thrust::complex by libcu++ #3507

Draft
wants to merge 1 commit into
base: main
Choose a base branch
from

Conversation

bernhardmgruber
Copy link
Contributor

No description provided.

Copy link

copy-pr-bot bot commented Jan 23, 2025

Auto-sync is disabled for draft pull requests in this repository. Workflows must be run manually.

Contributors can view more details about this message here.

@bernhardmgruber
Copy link
Contributor Author

Discussed with @miscco, apparently something similar was tried with #260 but abandoned. Later #454 tried to replace thrust complex partially, but it was reverted in #1497, because the numerical results did not work out. It seems cuda::std::complex still needs fixes.

@miscco
Copy link
Collaborator

miscco commented Jan 23, 2025

The issue is more with the math functions than with anything else

@davebayer
Copy link
Contributor

I think the biggest problem is that thrust::complex is trivially default constructible, but cuda::std::complex is not.

The replacement could lead to some unexpected behaviour. An example:

thrust::complex<float> alloc(size_t n)
{
  return new thrust::complex<float>[n];
}

The old implementation does not zero initialize the allocated array, but when implemented by cuda::std::complex, it does.

This can have some serious performance consequences, e. g. for NUMA systems where the memory will be first touched by a single thread.

@miscco
Copy link
Collaborator

miscco commented Feb 9, 2025

I think the biggest problem is that thrust::complex is trivially default constructible, but cuda::std::complex is not.

That is actually a quite important problem, as certain libraries require trivially_copyable types, for example rapids device_uvector

The long term solution that we want is a cuda::complex which effectively has all the functionality of cuda::std::complex but is just a plain aggregate. The issue there is that we would then need to duplicate all math functions because they need to take a cuda::std::complex and return a cuda::std::complex and cannot rely on e.g implicit conversion through inheritance

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
None yet
Projects
Status: In Progress
Development

Successfully merging this pull request may close these issues.

3 participants