Skip to content

Add environment overloads for DeviceTransform - #6204

Merged
bernhardmgruber merged 4 commits into
NVIDIA:mainfrom
bernhardmgruber:env_transform
Oct 15, 2025
Merged

bernhardmgruber merged 4 commits into
NVIDIA:mainfrom
bernhardmgruber:env_transform

Conversation

@bernhardmgruber

Copy link
Copy Markdown
Contributor

Fixes: #6203

@bernhardmgruber
bernhardmgruber requested a review from a team as a code owner October 13, 2025 13:51
@github-project-automation github-project-automation Bot moved this to Todo in CCCL Oct 13, 2025
@cccl-authenticator-app cccl-authenticator-app Bot moved this from Todo to In Review in CCCL Oct 13, 2025
Comment thread cub/cub/device/device_transform.cuh Resolved
Comment thread cub/cub/device/device_transform.cuh Resolved
@bernhardmgruber
bernhardmgruber enabled auto-merge (squash) October 13, 2025 14:30
@github-actions

This comment has been minimized.

@github-actions

Copy link
Copy Markdown
Contributor

🥳 CI Workflow Results

🟩 Finished in 3h 51m: Pass: 100%/75 | Total: 2d 02h | Max: 2h 21m | Hits: 75%/73732

See results here.

@bernhardmgruber
bernhardmgruber merged commit 66d5d89 into NVIDIA:main Oct 15, 2025
1 check passed
@github-project-automation github-project-automation Bot moved this from In Review to Done in CCCL Oct 15, 2025
@bernhardmgruber
bernhardmgruber deleted the env_transform branch October 16, 2025 10:02
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Jan 14, 2026
NVIDIA#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Jan 15, 2026
NVIDIA#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Jan 15, 2026
NVIDIA#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.
bernhardmgruber added a commit that referenced this pull request Jan 16, 2026
* Fix extracting CUDA stream in cub::DeviceTransform

#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.

* Properly use `get_stream` in device transform

We should not have a special overload for the `get_stream` overload but should ensure that we can pass types that are convertible to `::cudaStream_t` and extract that stream

I have expanded the get_stream CPO to accept a `::cudaStream_t __stream` so that we can work with those types

Co-authored-by: Michael Schellenberger Costa <miscco@nvidia.com>
github-actions Bot pushed a commit that referenced this pull request Jan 16, 2026
* Fix extracting CUDA stream in cub::DeviceTransform

#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.

* Properly use `get_stream` in device transform

We should not have a special overload for the `get_stream` overload but should ensure that we can pass types that are convertible to `::cudaStream_t` and extract that stream

I have expanded the get_stream CPO to accept a `::cudaStream_t __stream` so that we can work with those types

Co-authored-by: Michael Schellenberger Costa <miscco@nvidia.com>
(cherry picked from commit 31f8a13)
bernhardmgruber added a commit that referenced this pull request Jan 16, 2026
…7263)

* Fix extracting CUDA stream in cub::DeviceTransform

#6204 changed cub::DeviceTransform APIs from taking cudaStream_t to environments. Special handling preserved support for cudaStream_t. However, user-provided stream types with conversion operators to cudaStream_t were now queried as environments, failing to return a stream.

* Properly use `get_stream` in device transform

We should not have a special overload for the `get_stream` overload but should ensure that we can pass types that are convertible to `::cudaStream_t` and extract that stream

I have expanded the get_stream CPO to accept a `::cudaStream_t __stream` so that we can work with those types

* Fix issue with `__query_or`

* Avoid issue with `qury_or` CPO

---------

Co-authored-by: Michael Schellenberger Costa <miscco@nvidia.com>
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Mar 6, 2026
This is a fix for another bug introduced in NVIDIA#6204
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Mar 6, 2026
This is a fix for another bug introduced in NVIDIA#6204
bernhardmgruber added a commit that referenced this pull request Mar 12, 2026
This is a fix for another bug introduced in #6204
bernhardmgruber added a commit to bernhardmgruber/cccl that referenced this pull request Mar 12, 2026
wmaxey added a commit that referenced this pull request Mar 13, 2026
This is a fix for another bug introduced in #6204

Co-authored-by: Wesley Maxey <71408887+wmaxey@users.noreply.github.com>
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Labels

None yet

Projects

Archived in project

Development

Successfully merging this pull request may close these issues.

Add env-based cub::DeviceTransform

2 participants