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

[FEA]: cuda::span_collection #2938

Open
1 task done
davebayer opened this issue Nov 22, 2024 · 1 comment
Open
1 task done

[FEA]: cuda::span_collection #2938

davebayer opened this issue Nov 22, 2024 · 1 comment
Labels
feature request New feature or request.

Comments

@davebayer
Copy link
Contributor

davebayer commented Nov 22, 2024

Is this a duplicate?

Area

CUDA Experimental (cudax)

Is your feature request related to a problem? Please describe.

It is a common case that you need to pass views over multiple arrays to a function or kernel that have the exact same dimensions. It leads to code duplication and passing more arguments than necessary.

Describe the solution you'd like

I propose to implement a storage optimized container of views cuda::span_collection that could be used like this:

template<cuda::std::size_t extent>
void vector_add(cuda::span_collection<extent, const int, const int, int> params)
{
    for (auto [a, b, c] : params)
    {
        c = a + b;
    }
}

int main()
{
    static constexpr cuda::std::size_t size = 8;

    cuda::std::array<const int, size> a{1, 2, 3, 4, 5, 6, 7, 8};
    cuda::std::array<const int, size> b{1, 2, 3, 4, 5, 6, 7, 8};
    cuda::std::array<int, size>       c{};

    vector_add({size, a, b, c});
}

I tried to play with it a bit here https://godbolt.org/z/9WWhjKxfq

What are your thoughts on this idea?

Describe alternatives you've considered

No response

Additional context

No response

@davebayer davebayer added the feature request New feature or request. label Nov 22, 2024
@davebayer davebayer changed the title [FEA]: Span collection [FEA]: cuda::span_collection Nov 22, 2024
@bernhardmgruber
Copy link
Contributor

I believe we have something similar with thrust::zip_iterator, although we don't have a corresponding range type for it. C++23 has std::views::zip which should serve this purpose, and @miscco has somewhat of a ranges implementation here: #198.

Therefore, I don't see the necessity of introducing this feature, unless I misunderstood it.

Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment
Labels
feature request New feature or request.
Projects
None yet
Development

No branches or pull requests

2 participants