[cudax->libcu++] Move buffer type from cudax to libcu++ by pciolkosz · Pull Request #6627 · NVIDIA/cccl

pciolkosz · 2025-11-14T01:57:01Z

PR moving last bits of the buffer type and de-duplicating the __memory_accessibility enum

…udacxx

miscco · 2025-11-17T10:19:41Z

@@ -119,15 +119,6 @@ foreach (cudax_target IN LISTS cudax_TARGETS)

  cudax_add_catch2_test(test_target containers ${cudax_target}
    containers/uninitialized_buffer.cu


arent we moving uninitialized_buffer too

We don't really use it anywhere, I think it can stay experimental until we have a usage for it

What's the new replacement for thrust::device_vector<T>(n, thrust::no_init), which is used everywhere in benchmarks and tests?

cuda::make_buffer<T>(stream, cuda::device_default_memory_pool(cuda::devices[i]), n, cuda::no_init). We might consider a version of it in the future that does not take a stream

I think it would look much better if we could just do:

cuda::make_buffer<T>(stream, cuda::devices[i], n, cuda::no_init);

We could add a constructor from cuda::memory_location which would mean the default memory pool

…udacxx

davebayer · 2025-12-01T09:26:38Z

pre-commit.ci autofix

copy-pr-bot · 2025-12-01T09:28:21Z

This pull request requires additional validation before any workflows can run on NVIDIA's runners.

Pull request vetters can view their responsibilities here.

Contributors can view more details about this message here.

davebayer · 2025-12-01T09:52:15Z

pre-commit.ci autofix

davebayer · 2025-12-01T10:03:00Z

/ok to test c804cfe

copy-pr-bot · 2025-12-01T10:03:02Z

/ok to test c804cfe

@davebayer, there was an error processing your request: E2

See the following link for more information: https://docs.gha-runners.nvidia.com/cpr/e/2/

davebayer · 2025-12-01T10:03:26Z

/ok to test a804cfe

github-actions · 2025-12-01T19:44:32Z

🥳 CI Workflow Results

🟩 Finished in 9h 38m: Pass: 100%/122 | Total: 1d 21h | Max: 3h 01m | Hits: 99%/215255

See results here.

github-actions · 2025-12-01T21:36:15Z

Backport failed for branch/3.2.x, because it was unable to cherry-pick the commit(s).

Please cherry-pick the changes locally and resolve any conflicts.

git fetch origin branch/3.2.x
git worktree add -d .worktree/backport-6627-to-branch/3.2.x origin/branch/3.2.x
cd .worktree/backport-6627-to-branch/3.2.x
git switch --create backport-6627-to-branch/3.2.x
git cherry-pick -x 5dd8bdb567b92ddec5f81cf7f49a61d31dea41e7

@davebayer

* Move buffer type from cudax to libcu++ * Fix cudax compilation * Fix merge and review feedback * Waive buffer prologue include checks * Use unwrapped_begin in get_unsynchronized * ifdef problematic cases for now * Apply suggestion from @davebayer * [pre-commit.ci] auto code formatting * Apply suggestion from @davebayer --------- Co-authored-by: David Bayer <48736217+davebayer@users.noreply.github.com> Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com>

@davebayer

* Move buffer type from cudax to libcu++ * Fix cudax compilation * Fix merge and review feedback * Waive buffer prologue include checks * Use unwrapped_begin in get_unsynchronized * ifdef problematic cases for now * Apply suggestion from @davebayer * [pre-commit.ci] auto code formatting * Apply suggestion from @davebayer --------- Co-authored-by: David Bayer <48736217+davebayer@users.noreply.github.com> Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com>

davebayer · 2025-12-02T07:33:24Z

+// Temporary workaround to not trigger issues in CUB headers missing prologue
+#define _CCCL_WAIVE_PROLOGUE_INCLUDE_CHECK


What was the problem? I don't like that we are disabling this

CUB didn't properly include prologue headers and I didn't want to try to fix that as part of this PR

I understand but then there are these workarounds coming out of nowhere. I think it would be better to fix this problem in another PR before merging this one.

It seems that we will include more cub headers, so we need a systematic solution.

Fixed by #6844

I still think it would be nice to have an opt-out like that available for similar future cases, especially since the prologue thing is not essential, but I don't mind removing it

No, it is absolutely essential. All code in libcu++ should be guarded by the prologue/epilogue include. We do many important stuff there to prevent other macros breaking our code

But now we also include CUB in libcu++ and CUB doesn't use it, so I would be cautious to call it essential

And not only that, we suppress compiler warnings and other stuff. That's why the checks for the prologue/epilogue wrapping are so strict (and annoying)

CUB doesn't do many of the things we do in libcu++, but that doesn't mean we should relax requirements on what we do in libcu++

I am with Dave here, this is potentially breaking libcu++ which we must not do

@davebayer

* Move buffer type from cudax to libcu++ * Fix cudax compilation * Fix merge and review feedback * Waive buffer prologue include checks * Use unwrapped_begin in get_unsynchronized * ifdef problematic cases for now * Apply suggestion from @davebayer * [pre-commit.ci] auto code formatting * Apply suggestion from @davebayer --------- Co-authored-by: David Bayer <48736217+davebayer@users.noreply.github.com> Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com>

@davebayer

* Move buffer type from cudax to libcu++ * Fix cudax compilation * Fix merge and review feedback * Waive buffer prologue include checks * Use unwrapped_begin in get_unsynchronized * ifdef problematic cases for now * Apply suggestion from @davebayer * [pre-commit.ci] auto code formatting * Apply suggestion from @davebayer --------- Co-authored-by: David Bayer <48736217+davebayer@users.noreply.github.com> Co-authored-by: pre-commit-ci[bot] <66853113+pre-commit-ci[bot]@users.noreply.github.com>

Move buffer type from cudax to libcu++

5c556a4

pciolkosz requested review from a team as code owners November 14, 2025 01:57

pciolkosz requested a review from griwes November 14, 2025 01:57

github-project-automation Bot added this to CCCL Nov 14, 2025

github-project-automation Bot moved this to Todo in CCCL Nov 14, 2025

cccl-authenticator-app Bot moved this from Todo to In Review in CCCL Nov 14, 2025

This comment has been minimized.

Sign in to view

Merge remote-tracking branch 'upstream/main' into move_buffer_to_libc…

1248b3d

…udacxx

davebayer approved these changes Nov 14, 2025

View reviewed changes

Fix cudax compilation

4453614

pciolkosz requested a review from a team as a code owner November 14, 2025 23:09

pciolkosz requested a review from wmaxey November 14, 2025 23:09

This comment has been minimized.

Sign in to view

miscco reviewed Nov 17, 2025

View reviewed changes

davebayer reviewed Nov 19, 2025

View reviewed changes

Comment thread libcudacxx/include/cuda/__container/buffer.h

pciolkosz added 2 commits November 24, 2025 18:55

Merge remote-tracking branch 'upstream/main' into move_buffer_to_libc…

2c1109a

…udacxx

Fix merge and review feedback

630a00e

This comment has been minimized.

Sign in to view

pciolkosz added the backport branch/3.2.x label Nov 25, 2025

Merge branch 'main' into move_buffer_to_libcudacxx

88807a8

This comment has been minimized.

Sign in to view

Waive buffer prologue include checks

a519bd0

This comment has been minimized.

Sign in to view

Use unwrapped_begin in get_unsynchronized

9274d13

This comment has been minimized.

Sign in to view

pciolkosz added 2 commits November 30, 2025 22:10

ifdef problematic cases for now

64e6241

Merge branch 'main' into move_buffer_to_libcudacxx

b4c0d74

This comment has been minimized.

Sign in to view

davebayer reviewed Dec 1, 2025

View reviewed changes

Comment thread libcudacxx/include/cuda/__container/heterogeneous_iterator.h Outdated

Apply suggestion from @davebayer

dc27d41

[pre-commit.ci] auto code formatting

f0ae25b

davebayer reviewed Dec 1, 2025

View reviewed changes

Comment thread libcudacxx/test/libcudacxx/cuda/containers/buffer/properties.cu Outdated

Apply suggestion from @davebayer

a804cfe

pciolkosz merged commit 5dd8bdb into NVIDIA:main Dec 1, 2025
135 of 136 checks passed

github-project-automation Bot moved this from In Review to Done in CCCL Dec 1, 2025

davebayer reviewed Dec 2, 2025

View reviewed changes

harrism mentioned this pull request Aug 5, 2026

cuda::buffer::set_stream documents synchronization but calls set_stream_unsynchronized #10649

Closed

		@@ -119,15 +119,6 @@ foreach (cudax_target IN LISTS cudax_TARGETS)

		cudax_add_catch2_test(test_target containers ${cudax_target}
		containers/uninitialized_buffer.cu

		// Temporary workaround to not trigger issues in CUB headers missing prologue
		#define _CCCL_WAIVE_PROLOGUE_INCLUDE_CHECK

Conversation

pciolkosz commented Nov 14, 2025

Uh oh!

This comment has been minimized.

This comment has been minimized.

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

bernhardmgruber Nov 25, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Choose a reason for hiding this comment

Uh oh!

pciolkosz Nov 25, 2025 • edited Loading Uh oh! There was an error while loading. Please reload this page.

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

Uh oh!

This comment has been minimized.

This comment has been minimized.

This comment has been minimized.

This comment has been minimized.

This comment has been minimized.

Uh oh!

davebayer commented Dec 1, 2025

Uh oh!

copy-pr-bot Bot commented Dec 1, 2025

Uh oh!

Uh oh!

davebayer commented Dec 1, 2025

Uh oh!

davebayer commented Dec 1, 2025

Uh oh!

copy-pr-bot Bot commented Dec 1, 2025

Uh oh!

davebayer commented Dec 1, 2025

Uh oh!

github-actions Bot commented Dec 1, 2025

🥳 CI Workflow Results

🟩 Finished in 9h 38m: Pass: 100%/122 | Total: 1d 21h | Max: 3h 01m | Hits: 99%/215255

Uh oh!

Uh oh!

github-actions Bot commented Dec 1, 2025

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Choose a reason for hiding this comment

Uh oh!

Reviewers

Assignees

Labels

Projects

Milestone

Development

Uh oh!

4 participants

bernhardmgruber Nov 25, 2025 •

edited

Loading

pciolkosz Nov 25, 2025 •

edited

Loading