From d1716e11afdda3a0c4ef98452cd653c1e9fe0996 Mon Sep 17 00:00:00 2001 From: gonidelis Date: Fri, 13 Mar 2026 19:20:12 -0700 Subject: [PATCH] Don't use cudaStream_t legacy --- ...atch2_test_device_run_length_encode_env.cu | 22 +++++++++---------- 1 file changed, 10 insertions(+), 12 deletions(-) diff --git a/cub/test/catch2_test_device_run_length_encode_env.cu b/cub/test/catch2_test_device_run_length_encode_env.cu index bb163648ae5..50e79bbef00 100644 --- a/cub/test/catch2_test_device_run_length_encode_env.cu +++ b/cub/test/catch2_test_device_run_length_encode_env.cu @@ -11,7 +11,9 @@ struct stream_registry_factory_t; #include +#include #include +#include #include "catch2_test_env_launch_helper.h" @@ -152,8 +154,7 @@ TEST_CASE("DeviceRunLengthEncode::Encode uses custom stream", "[run_length_encod auto d_num_runs_out = c2h::device_vector(1); int num_items = static_cast(d_in.size()); - cudaStream_t custom_stream; - REQUIRE(cudaSuccess == cudaStreamCreate(&custom_stream)); + cuda::stream custom_stream{cuda::devices[0]}; size_t expected_bytes_allocated{}; REQUIRE( @@ -167,13 +168,14 @@ TEST_CASE("DeviceRunLengthEncode::Encode uses custom stream", "[run_length_encod d_num_runs_out.begin(), num_items)); - auto stream_prop = stdexec::prop{cuda::get_stream_t{}, cuda::stream_ref{custom_stream}}; + cuda::stream_ref stream_ref{custom_stream}; + auto stream_prop = stdexec::prop{cuda::get_stream_t{}, stream_ref}; auto env = stdexec::env{stream_prop, expected_allocation_size(expected_bytes_allocated)}; run_length_encode_env( d_in.begin(), d_unique_out.begin(), d_counts_out.begin(), d_num_runs_out.begin(), num_items, env); - REQUIRE(cudaSuccess == cudaStreamSynchronize(custom_stream)); + custom_stream.sync(); c2h::device_vector expected_unique{0, 2, 9, 5, 8}; c2h::device_vector expected_counts{1, 2, 1, 3, 1}; @@ -184,8 +186,6 @@ TEST_CASE("DeviceRunLengthEncode::Encode uses custom stream", "[run_length_encod d_counts_out.resize(d_num_runs_out[0]); REQUIRE(d_unique_out == expected_unique); REQUIRE(d_counts_out == expected_counts); - - REQUIRE(cudaSuccess == cudaStreamDestroy(custom_stream)); } TEST_CASE("DeviceRunLengthEncode::NonTrivialRuns uses custom stream", "[run_length_encode][device]") @@ -196,8 +196,7 @@ TEST_CASE("DeviceRunLengthEncode::NonTrivialRuns uses custom stream", "[run_leng auto d_num_runs_out = c2h::device_vector(1); int num_items = static_cast(d_in.size()); - cudaStream_t custom_stream; - REQUIRE(cudaSuccess == cudaStreamCreate(&custom_stream)); + cuda::stream custom_stream{cuda::devices[0]}; size_t expected_bytes_allocated{}; REQUIRE( @@ -211,13 +210,14 @@ TEST_CASE("DeviceRunLengthEncode::NonTrivialRuns uses custom stream", "[run_leng d_num_runs_out.begin(), num_items)); - auto stream_prop = stdexec::prop{cuda::get_stream_t{}, cuda::stream_ref{custom_stream}}; + cuda::stream_ref stream_ref{custom_stream}; + auto stream_prop = stdexec::prop{cuda::get_stream_t{}, stream_ref}; auto env = stdexec::env{stream_prop, expected_allocation_size(expected_bytes_allocated)}; non_trivial_runs_env( d_in.begin(), d_offsets_out.begin(), d_lengths_out.begin(), d_num_runs_out.begin(), num_items, env); - REQUIRE(cudaSuccess == cudaStreamSynchronize(custom_stream)); + custom_stream.sync(); c2h::device_vector expected_offsets{1, 4}; c2h::device_vector expected_lengths{2, 3}; @@ -228,6 +228,4 @@ TEST_CASE("DeviceRunLengthEncode::NonTrivialRuns uses custom stream", "[run_leng d_lengths_out.resize(d_num_runs_out[0]); REQUIRE(d_offsets_out == expected_offsets); REQUIRE(d_lengths_out == expected_lengths); - - REQUIRE(cudaSuccess == cudaStreamDestroy(custom_stream)); }