Deprecate <cuda/stream_ref> header (#6266)

davebayer · web-flow · commit c99060c19e91 · 2025-10-16T14:44:39.000Z
diff --git a/cub/cub/device/device_reduce.cuh b/cub/cub/device/device_reduce.cuh
@@ -60,6 +60,7 @@
 #include <cuda/__iterator/tabulate_output_iterator.h>
 #include <cuda/__memory_resource/get_memory_resource.h>
 #include <cuda/__stream/get_stream.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__execution/env.h>
 #include <cuda/std/__functional/identity.h>
 #include <cuda/std/__functional/invoke.h>
@@ -69,7 +70,6 @@
 #include <cuda/std/__type_traits/is_same.h>
 #include <cuda/std/cstdint>
 #include <cuda/std/limits>
-#include <cuda/stream_ref>
 
 CUB_NAMESPACE_BEGIN
 
diff --git a/cudax/examples/vector.cuh b/cudax/examples/vector.cuh
@@ -24,9 +24,9 @@
 #include <thrust/device_vector.h>
 #include <thrust/host_vector.h>
 
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__type_traits/maybe_const.h>
 #include <cuda/std/span>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__detail/utility.cuh>
 #include <cuda/experimental/__launch/param_kind.cuh>
diff --git a/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh b/cudax/include/cuda/experimental/__container/uninitialized_async_buffer.cuh
@@ -22,6 +22,7 @@
 #endif // no system header
 
 #include <cuda/__memory_resource/properties.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__memory/addressof.h>
 #include <cuda/std/__memory/align.h>
 #include <cuda/std/__new/launder.h>
@@ -30,7 +31,6 @@
 #include <cuda/std/__utility/move.h>
 #include <cuda/std/__utility/swap.h>
 #include <cuda/std/span>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__memory_resource/any_resource.cuh>
 
diff --git a/cudax/include/cuda/experimental/__launch/host_launch.cuh b/cudax/include/cuda/experimental/__launch/host_launch.cuh
@@ -21,11 +21,11 @@
 #  pragma system_header
 #endif // no system header
 
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__functional/reference_wrapper.h>
 #include <cuda/std/__type_traits/decay.h>
 #include <cuda/std/__utility/forward.h>
 #include <cuda/std/tuple>
-#include <cuda/stream_ref>
 
 #include <cuda/std/__cccl/prologue.h>
 
diff --git a/cudax/include/cuda/experimental/__launch/launch.cuh b/cudax/include/cuda/experimental/__launch/launch.cuh
@@ -22,13 +22,13 @@
 #endif // no system header
 
 #include <cuda/__driver/driver_api.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__exception/cuda_error.h>
 #include <cuda/std/__type_traits/is_function.h>
 #include <cuda/std/__type_traits/is_pointer.h>
 #include <cuda/std/__type_traits/type_identity.h>
 #include <cuda/std/__utility/forward.h>
 #include <cuda/std/__utility/pod_tuple.h>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__execution/completion_signatures.cuh>
 #include <cuda/experimental/__execution/cpos.cuh>
diff --git a/cudax/include/cuda/experimental/__memory_resource/device_memory_pool.cuh b/cudax/include/cuda/experimental/__memory_resource/device_memory_pool.cuh
@@ -28,10 +28,10 @@
 
 #include <cuda/__memory_resource/get_property.h>
 #include <cuda/__memory_resource/properties.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__cuda/api_wrapper.h>
 #include <cuda/std/__new_>
 #include <cuda/std/span>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__memory_resource/memory_pool_base.cuh>
 #include <cuda/experimental/__stream/stream.cuh>
diff --git a/cudax/include/cuda/experimental/__memory_resource/memory_resource_base.cuh b/cudax/include/cuda/experimental/__memory_resource/memory_resource_base.cuh
@@ -27,11 +27,10 @@
 #endif // _CCCL_CUDA_COMPILER(CLANG)
 
 #include <cuda/__device/device_ref.h>
-#include <cuda/__memory_resource/properties.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__concepts/concept_macros.h>
 #include <cuda/std/__cuda/api_wrapper.h>
 #include <cuda/std/cstddef>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__memory_resource/any_resource.cuh>
 #include <cuda/experimental/__memory_resource/memory_pool_base.cuh>
diff --git a/cudax/include/cuda/experimental/__stream/device_transform.cuh b/cudax/include/cuda/experimental/__stream/device_transform.cuh
@@ -21,6 +21,7 @@
 #  pragma system_header
 #endif // no system header
 
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__memory/addressof.h>
 #include <cuda/std/__memory/construct_at.h>
 #include <cuda/std/__new/launder.h>
@@ -30,7 +31,6 @@
 #include <cuda/std/__utility/declval.h>
 #include <cuda/std/__utility/forward.h>
 #include <cuda/std/__utility/move.h>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__detail/type_traits.cuh>
 #include <cuda/experimental/__detail/utility.cuh>
diff --git a/cudax/include/cuda/experimental/__stream/stream_ref.cuh b/cudax/include/cuda/experimental/__stream/stream_ref.cuh
@@ -22,8 +22,8 @@
 
 #include <cuda/__device/all_devices.h>
 #include <cuda/__event/timed_event.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__cuda/api_wrapper.h>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/__device/logical_device.cuh>
 #include <cuda/experimental/__execution/completion_behavior.cuh>
diff --git a/cudax/test/common/utility.cuh b/cudax/test/common/utility.cuh
@@ -14,10 +14,10 @@
 #include <cuda_runtime_api.h>
 // cuda_runtime_api needs to come first
 
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/atomic>
 #include <cuda/std/__cuda/api_wrapper.h>
 #include <cuda/std/utility>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/launch.cuh>
 
diff --git a/cudax/test/containers/async_buffer/test_resources.h b/cudax/test/containers/async_buffer/test_resources.h
@@ -11,9 +11,9 @@
 #ifndef CUDAX_TEST_CONTAINER_VECTOR_TEST_RESOURCES_H
 #define CUDAX_TEST_CONTAINER_VECTOR_TEST_RESOURCES_H
 
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/memory_resource>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
 
 #include <cuda/experimental/memory_resource.cuh>
 
diff --git a/cudax/test/containers/uninitialized_async_buffer.cu b/cudax/test/containers/uninitialized_async_buffer.cu
@@ -17,7 +17,7 @@
 #include <cuda/std/span>
 #include <cuda/std/type_traits>
 #include <cuda/std/utility>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/container.cuh>
 #include <cuda/experimental/memory_resource.cuh>
diff --git a/cudax/test/memory_resource/device_memory_resource.cu b/cudax/test/memory_resource/device_memory_resource.cu
@@ -10,7 +10,7 @@
 
 #include <cuda/std/cstdint>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/memory_resource.cuh>
 
diff --git a/cudax/test/memory_resource/managed_memory_resource.cu b/cudax/test/memory_resource/managed_memory_resource.cu
@@ -10,7 +10,7 @@
 
 #include <cuda/std/cstdint>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/memory_resource.cuh>
 #include <cuda/experimental/stream.cuh>
diff --git a/cudax/test/memory_resource/memory_pools.cu b/cudax/test/memory_resource/memory_pools.cu
@@ -10,7 +10,7 @@
 
 #include <cuda/std/cstdint>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/launch.cuh>
 #include <cuda/experimental/memory_resource.cuh>
diff --git a/cudax/test/memory_resource/pinned_memory_resource.cu b/cudax/test/memory_resource/pinned_memory_resource.cu
@@ -10,7 +10,7 @@
 
 #include <cuda/std/cstdint>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/memory_resource.cuh>
 #include <cuda/experimental/stream.cuh>
diff --git a/docs/libcudacxx/extended_api/memory_resource.rst b/docs/libcudacxx/extended_api/memory_resource.rst
@@ -46,4 +46,4 @@ that are e.g. accessible on device, or whether it can utilize other allocation m
 
 To better support asynchronous CUDA `stream-ordered allocations <https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#stream-ordered-memory-allocator>`__
 libcu++ provides :ref:`cuda::stream_ref <libcudacxx-extended-api-streams-stream-ref>` as a wrapper around
-``cudaStream_t``. The definition of ``cuda::stream_ref`` can be found in the ``<cuda/stream_ref>`` header.
+``cudaStream_t``. The definition of ``cuda::stream_ref`` can be found in the ``<cuda/stream>`` header.
diff --git a/examples/cudax/vector_add/vector.cuh b/examples/cudax/vector_add/vector.cuh
@@ -26,7 +26,7 @@
 
 #include <cuda/std/__type_traits/maybe_const.h>
 #include <cuda/std/span>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <cuda/experimental/__detail/utility.cuh>
 #include <cuda/experimental/__launch/param_kind.cuh>
diff --git a/libcudacxx/cmake/LibcudacxxBuildCompilerTargets.cmake b/libcudacxx/cmake/LibcudacxxBuildCompilerTargets.cmake
@@ -22,6 +22,7 @@ function(libcudacxx_build_compiler_targets)
   list(APPEND cxx_compile_definitions "CCCL_IGNORE_DEPRECATED_CPP_DIALECT")
   list(APPEND cxx_compile_definitions "CCCL_ENABLE_OPTIONAL_REF")
   list(APPEND cxx_compile_definitions "CCCL_IGNORE_DEPRECATED_DISCARD_MEMORY_HEADER")
+  list(APPEND cxx_compile_definitions "CCCL_IGNORE_DEPRECATED_STREAM_REF_HEADER")
 
   cccl_build_compiler_interface(libcudacxx.compiler_interface
     "${cuda_compile_options}"
diff --git a/libcudacxx/include/cuda/__memory_resource/get_memory_resource.h b/libcudacxx/include/cuda/__memory_resource/get_memory_resource.h
@@ -23,11 +23,11 @@
 
 #include <cuda/__memory_resource/properties.h>
 #include <cuda/__memory_resource/resource.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__concepts/equality_comparable.h>
 #include <cuda/std/__execution/env.h>
 #include <cuda/std/__type_traits/is_same.h>
 #include <cuda/std/__type_traits/remove_cvref.h>
-#include <cuda/stream_ref>
 
 #include <cuda/std/__cccl/prologue.h>
 
diff --git a/libcudacxx/include/cuda/__memory_resource/resource.h b/libcudacxx/include/cuda/__memory_resource/resource.h
@@ -22,14 +22,14 @@
 #endif // no system header
 
 #include <cuda/__memory_resource/get_property.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/std/__concepts/concept_macros.h>
 #include <cuda/std/__concepts/convertible_to.h>
 #include <cuda/std/__concepts/equality_comparable.h>
 #include <cuda/std/__concepts/same_as.h>
 #include <cuda/std/__tuple_dir/sfinae_helpers.h>
 #include <cuda/std/__type_traits/decay.h>
 #include <cuda/std/__type_traits/fold.h>
-#include <cuda/stream_ref>
 
 #include <cuda/std/__cccl/prologue.h>
 
diff --git a/libcudacxx/include/cuda/__memory_resource/resource_ref.h b/libcudacxx/include/cuda/__memory_resource/resource_ref.h
@@ -26,6 +26,7 @@
 #  include <cuda/__memory_resource/get_property.h>
 #  include <cuda/__memory_resource/properties.h>
 #  include <cuda/__memory_resource/resource.h>
+#  include <cuda/__stream/stream_ref.h>
 #  include <cuda/std/__concepts/concept_macros.h>
 #  include <cuda/std/__memory/addressof.h>
 #  include <cuda/std/__type_traits/is_base_of.h>
@@ -34,7 +35,6 @@
 #  include <cuda/std/__utility/exchange.h>
 #  include <cuda/std/__utility/move.h>
 #  include <cuda/std/cstddef>
-#  include <cuda/stream_ref>
 
 #  include <cuda/std/__cccl/prologue.h>
 
diff --git a/libcudacxx/include/cuda/stream_ref b/libcudacxx/include/cuda/stream_ref
@@ -48,6 +48,11 @@ private:
 #  pragma system_header
 #endif // no system header
 
+#if !defined(CCCL_IGNORE_DEPRECATED_STREAM_REF_HEADER)
+_CCCL_WARNING("The <cuda/stream_ref> header is deprecated and will be removed in the next major release. The \
+contents were moved to <cuda/stream>. To ignore this message, define CCCL_IGNORE_DEPRECATED_STREAM_REF_HEADER.")
+#endif // !CCCL_IGNORE_DEPRECATED_STREAM_REF_HEADER
+
 #include <cuda/__stream/get_stream.h>
 #include <cuda/__stream/stream_ref.h>
 
diff --git a/libcudacxx/test/libcudacxx/cuda/ccclrt/common/utility.cuh b/libcudacxx/test/libcudacxx/cuda/ccclrt/common/utility.cuh
@@ -15,10 +15,10 @@
 // cuda_runtime_api needs to come first
 
 #include <cuda/__runtime/ensure_current_context.h>
+#include <cuda/__stream/stream_ref.h>
 #include <cuda/atomic>
 #include <cuda/std/__cuda/api_wrapper.h>
 #include <cuda/std/utility>
-#include <cuda/stream_ref>
 
 #include <new> // IWYU pragma: keep (needed for placement new)
 
diff --git a/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/equality.fail.cpp b/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/equality.fail.cpp
@@ -15,7 +15,7 @@
 #include <cuda/memory_resource>
 #include <cuda/std/cassert>
 #include <cuda/std/cstdint>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include "types.h"
 
diff --git a/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/equality.pass.cpp b/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/equality.pass.cpp
@@ -15,7 +15,7 @@
 #include <cuda/memory_resource>
 #include <cuda/std/cassert>
 #include <cuda/std/cstdint>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include "types.h"
 
diff --git a/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/properties.pass.cpp b/libcudacxx/test/libcudacxx/cuda/memory_resource/async_resource_ref/properties.pass.cpp
@@ -15,7 +15,7 @@
 #include <cuda/memory_resource>
 #include <cuda/std/cassert>
 #include <cuda/std/cstdint>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include "test_macros.h"
 #include "types.h"
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/get_stream.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/get_stream.pass.cpp
@@ -10,7 +10,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 __host__ __device__ void test()
 {
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.constructor.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.constructor.pass.cpp
@@ -11,7 +11,7 @@
 
 #include <cuda/std/cassert>
 #include <cuda/std/type_traits>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 static_assert(cuda::std::is_default_constructible<cuda::stream_ref>::value, "");
 static_assert(!cuda::std::is_constructible<cuda::stream_ref, int>::value, "");
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.equality.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.equality.pass.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 __host__ __device__ void test()
 {
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.get.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.get.pass.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 __host__ __device__ void test()
 {
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.query.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.query.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 __host__ __device__ void test()
 {
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.ready.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.ready.pass.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include "test_macros.h"
 
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.sync.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.sync.pass.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <atomic>
 #include <chrono>
diff --git a/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.wait.pass.cpp b/libcudacxx/test/libcudacxx/cuda/stream_ref/stream_ref.wait.pass.cpp
@@ -11,7 +11,7 @@
 // UNSUPPORTED: nvrtc
 
 #include <cuda/std/cassert>
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <atomic>
 #include <chrono>
diff --git a/thrust/testing/cuda/stream_wrapper.cu b/thrust/testing/cuda/stream_wrapper.cu
@@ -2,7 +2,7 @@
 #include <thrust/iterator/constant_iterator.h>
 #include <thrust/reduce.h>
 
-#include <cuda/stream_ref>
+#include <cuda/stream>
 
 #include <unittest/unittest.h>
 
