diff --git a/cpp/bench/sg/arima_loglikelihood.cu b/cpp/bench/sg/arima_loglikelihood.cu index c779942750..2be184af6a 100644 --- a/cpp/bench/sg/arima_loglikelihood.cu +++ b/cpp/bench/sg/arima_loglikelihood.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,6 +14,7 @@ #include +#include #include #include #include @@ -33,9 +34,9 @@ class ArimaLoglikelihood : public TsFixtureRandom { ArimaLoglikelihood(const std::string& name, const ArimaParams& p) : TsFixtureRandom(name, p.data), order(p.order), - param(0, rmm::cuda_stream_default), - loglike(0, rmm::cuda_stream_default), - temp_mem(0, rmm::cuda_stream_default) + param(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}), + loglike(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}), + temp_mem(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}) { } @@ -45,7 +46,7 @@ class ArimaLoglikelihood : public TsFixtureRandom { using MLCommon::Bench::CudaEventTimer; auto& handle = *this->handle; - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto counting = thrust::make_counting_iterator(0); // Generate random parameters diff --git a/cpp/bench/sg/benchmark.cuh b/cpp/bench/sg/benchmark.cuh index b3d89f67b8..44237050ea 100644 --- a/cpp/bench/sg/benchmark.cuh +++ b/cpp/bench/sg/benchmark.cuh @@ -14,7 +14,7 @@ #include #include -#include +#include #include diff --git a/cpp/bench/sg/dataset.cuh b/cpp/bench/sg/dataset.cuh index d411816486..920728d845 100644 --- a/cpp/bench/sg/dataset.cuh +++ b/cpp/bench/sg/dataset.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,6 +14,8 @@ #include #include +#include + #include #include #include @@ -65,7 +67,11 @@ struct RegressionParams { */ template struct Dataset { - Dataset() : X(0, rmm::cuda_stream_default), y(0, rmm::cuda_stream_default) {} + Dataset() + : X(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}), + y(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}) + { + } /** input data */ rmm::device_uvector X; /** labels or output associated with each row of input data */ @@ -97,7 +103,7 @@ struct Dataset { void blobs(const raft::handle_t& handle, const DatasetParams& p, const BlobsParams& b) { const auto& handle_impl = handle; - auto stream = handle_impl.get_stream(); + auto stream = handle_impl.get_stream().get(); auto cublas_handle = handle_impl.get_cublas_handle(); // Make blobs will generate labels of type IdxT which has to be an integer @@ -139,7 +145,7 @@ struct Dataset { { ASSERT(!isClassification(), "make_regression: is only for regression problems!"); const auto& handle_impl = handle; - auto stream = handle_impl.get_stream(); + auto stream = handle_impl.get_stream().get(); auto cublas_handle = handle_impl.get_cublas_handle(); auto cusolver_handle = handle_impl.get_cusolver_dn_handle(); diff --git a/cpp/bench/sg/dataset_ts.cuh b/cpp/bench/sg/dataset_ts.cuh index d34b7dbd74..824fa7482c 100644 --- a/cpp/bench/sg/dataset_ts.cuh +++ b/cpp/bench/sg/dataset_ts.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2022, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -10,6 +10,8 @@ #include #include +#include + namespace ML { namespace Bench { @@ -26,7 +28,7 @@ struct TimeSeriesParams { */ template struct TimeSeriesDataset { - TimeSeriesDataset() : X(0, rmm::cuda_stream_default) {} + TimeSeriesDataset() : X(0, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}) {} /** input data */ rmm::device_uvector X; @@ -44,7 +46,7 @@ struct TimeSeriesDataset { DataT sigma = 1) { raft::random::Rng gpu_gen(p.seed, raft::random::GenPhilox); - gpu_gen.normal(X.data(), p.batch_size * p.n_obs, mu, sigma, handle.get_stream()); + gpu_gen.normal(X.data(), p.batch_size * p.n_obs, mu, sigma, handle.get_stream().get()); } }; diff --git a/cpp/src/arima/batched_arima.cu b/cpp/src/arima/batched_arima.cu index 1973aa35aa..3e69b3da67 100644 --- a/cpp/src/arima/batched_arima.cu +++ b/cpp/src/arima/batched_arima.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -43,7 +43,7 @@ void pack(raft::handle_t& handle, int batch_size, double* param_vec) { - const auto stream = handle.get_stream(); + const auto stream = handle.get_stream().get(); params.pack(order, batch_size, param_vec, stream); } @@ -53,7 +53,7 @@ void unpack(raft::handle_t& handle, int batch_size, const double* param_vec) { - const auto stream = handle.get_stream(); + const auto stream = handle.get_stream().get(); params.unpack(order, batch_size, param_vec, stream); } @@ -64,7 +64,7 @@ void batched_diff(raft::handle_t& handle, int n_obs, const ARIMAOrder& order) { - const auto stream = handle.get_stream(); + const auto stream = handle.get_stream().get(); MLCommon::TimeSeries::prepare_data( d_y_diff, d_y, batch_size, n_obs, order.d, order.D, order.s, stream); } @@ -80,7 +80,7 @@ struct is_missing { bool detect_missing(raft::handle_t& handle, const double* d_y, int n_elem) { return thrust::any_of( - thrust::cuda::par.on(handle.get_stream()), d_y, d_y + n_elem, is_missing()); + thrust::cuda::par.on(handle.get_stream().get()), d_y, d_y + n_elem, is_missing()); } void predict(raft::handle_t& handle, @@ -101,7 +101,7 @@ void predict(raft::handle_t& handle, double* d_upper) { raft::common::nvtx::range fun_scope(__func__); - const auto stream = handle.get_stream(); + const auto stream = handle.get_stream().get(); bool diff = order.need_diff() && pre_diff && level == 0; int num_steps = std::max(end - n_obs, 0); @@ -356,7 +356,7 @@ void conditional_sum_of_squares(raft::handle_t& handle, int truncate) { raft::common::nvtx::range fun_scope(__func__); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int n_phi = order.n_phi(); int n_theta = order.n_theta(); @@ -412,7 +412,7 @@ void batched_loglike(raft::handle_t& handle, { raft::common::nvtx::range fun_scope(__func__); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); double* d_pred = arima_mem.pred; @@ -485,7 +485,7 @@ void batched_loglike(raft::handle_t& handle, raft::common::nvtx::range fun_scope(__func__); // unpack parameters - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); ARIMAParams params = {arima_mem.params_mu, arima_mem.params_beta, @@ -527,7 +527,7 @@ void batched_loglike_grad(raft::handle_t& handle, int truncate) { raft::common::nvtx::range fun_scope(__func__); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto counting = thrust::make_counting_iterator(0); int N = order.complexity(); @@ -601,7 +601,7 @@ void information_criterion(raft::handle_t& handle, int ic_type) { raft::common::nvtx::range fun_scope(__func__); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); /* Compute log-likelihood in d_ic */ batched_loglike( @@ -674,7 +674,7 @@ void _arma_least_squares(raft::handle_t& handle, double* d_mu = nullptr) { const auto& handle_impl = handle; - auto stream = handle_impl.get_stream(); + auto stream = handle_impl.get_stream().get(); auto cublas_handle = handle_impl.get_cublas_handle(); auto counting = thrust::make_counting_iterator(0); @@ -956,7 +956,7 @@ void estimate_x0(raft::handle_t& handle, { raft::common::nvtx::range fun_scope(__func__); const auto& handle_impl = handle; - auto stream = handle_impl.get_stream(); + auto stream = handle_impl.get_stream().get(); auto cublas_handle = handle_impl.get_cublas_handle(); /// TODO: solve exogenous coefficients with only valid rows instead of interpolation? diff --git a/cpp/src/arima/batched_kalman.cu b/cpp/src/arima/batched_kalman.cu index 435a12fd77..29642bda49 100644 --- a/cpp/src/arima/batched_kalman.cu +++ b/cpp/src/arima/batched_kalman.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -850,7 +850,7 @@ void _lyapunov_wrapper(raft::handle_t& handle, int r) { if (r <= 5) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto cublasHandle = handle.get_cublas_handle(); int batch_size = ML::narrow_cast(A.batches()); int r2 = r * r; @@ -909,7 +909,7 @@ void _batched_kalman_filter(raft::handle_t& handle, double* d_upper) { const size_t batch_size = Zb.batches(); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto cublasHandle = handle.get_cublas_handle(); auto counting = thrust::make_counting_iterator(0); @@ -1152,7 +1152,7 @@ void init_batched_kalman_matrices(raft::handle_t& handle, { raft::common::nvtx::range fun_scope(__func__); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); // Note: Z is unused yet but kept to avoid reintroducing it later when // adding support for exogeneous variables @@ -1265,7 +1265,7 @@ void batched_kalman_filter(raft::handle_t& handle, raft::common::nvtx::range fun_scope(__func__); auto cublasHandle = handle.get_cublas_handle(); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); // see (3.18) in TSA by D&K int rd = order.rd(); @@ -1324,7 +1324,7 @@ void batched_jones_transform(raft::handle_t& handle, double* h_Tparams) { int N = order.complexity(); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); double* d_params = arima_mem.d_params; double* d_Tparams = arima_mem.d_Tparams; ARIMAParams params = {arima_mem.params_mu, diff --git a/cpp/src/datasets/make_arima.cu b/cpp/src/datasets/make_arima.cu index fc2fbcdee6..1554e4818f 100644 --- a/cpp/src/datasets/make_arima.cu +++ b/cpp/src/datasets/make_arima.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -23,7 +23,7 @@ inline void make_arima_helper(const raft::handle_t& handle, DataT intercept_scale, uint64_t seed) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); MLCommon::Random::make_arima( out, batch_size, n_obs, order, stream, scale, noise_scale, intercept_scale, seed); diff --git a/cpp/src/datasets/make_blobs.cu b/cpp/src/datasets/make_blobs.cu index 29cee45564..5ad39c453f 100644 --- a/cpp/src/datasets/make_blobs.cu +++ b/cpp/src/datasets/make_blobs.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -31,7 +31,7 @@ void make_blobs(const raft::handle_t& handle, n_rows, n_cols, n_clusters, - handle.get_stream(), + handle.get_stream().get(), row_major, centers, cluster_std, @@ -62,7 +62,7 @@ void make_blobs(const raft::handle_t& handle, n_rows, n_cols, n_clusters, - handle.get_stream(), + handle.get_stream().get(), row_major, centers, cluster_std, @@ -93,7 +93,7 @@ void make_blobs(const raft::handle_t& handle, n_rows, n_cols, n_clusters, - handle.get_stream(), + handle.get_stream().get(), row_major, centers, cluster_std, @@ -124,7 +124,7 @@ void make_blobs(const raft::handle_t& handle, n_rows, n_cols, n_clusters, - handle.get_stream(), + handle.get_stream().get(), row_major, centers, cluster_std, diff --git a/cpp/src/datasets/make_regression.cu b/cpp/src/datasets/make_regression.cu index 040e59107c..69e67f3ce3 100644 --- a/cpp/src/datasets/make_regression.cu +++ b/cpp/src/datasets/make_regression.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -28,7 +28,7 @@ void make_regression_helper(const raft::handle_t& handle, uint64_t seed) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); cublasHandle_t cublas_handle = handle_impl.get_cublas_handle(); cusolverDnHandle_t cusolver_handle = handle_impl.get_cusolver_dn_handle(); diff --git a/cpp/src/dbscan/dbscan.cu b/cpp/src/dbscan/dbscan.cu index 00ce78f8e9..34613a8e4d 100644 --- a/cpp/src/dbscan/dbscan.cu +++ b/cpp/src/dbscan/dbscan.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -41,7 +41,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); else dbscanFitImpl(handle, @@ -56,7 +56,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); } @@ -88,7 +88,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); else dbscanFitImpl(handle, @@ -103,7 +103,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); } @@ -135,7 +135,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); else dbscanFitImpl(handle, @@ -150,7 +150,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); } @@ -182,7 +182,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); else dbscanFitImpl(handle, @@ -197,7 +197,7 @@ void fit(const raft::handle_t& handle, sample_weight, max_bytes_per_batch, eps_nn_method, - handle.get_stream(), + handle.get_stream().get(), verbosity); } diff --git a/cpp/src/decisiontree/batched-levelalgo/quantiles.cuh b/cpp/src/decisiontree/batched-levelalgo/quantiles.cuh index 8d7b075cb3..d357134451 100644 --- a/cpp/src/decisiontree/batched-levelalgo/quantiles.cuh +++ b/cpp/src/decisiontree/batched-levelalgo/quantiles.cuh @@ -155,7 +155,7 @@ CUML_EXPORT QuantileResult computeQuantiles(const raft::handle_t& handle, bool row_major = false) { raft::common::nvtx::push_range("computeQuantiles"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); bool distributed = raft::resource::comms_initialized(handle) && handle.get_comms().get_size() > 1; RAFT_EXPECTS(max_n_bins > 0, "max_n_bins must be positive"); diff --git a/cpp/src/explainer/kernel_shap.cu b/cpp/src/explainer/kernel_shap.cu index 088609f1a3..2516c08203 100644 --- a/cpp/src/explainer/kernel_shap.cu +++ b/cpp/src/explainer/kernel_shap.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -172,7 +172,7 @@ void kernel_dataset_impl(const raft::handle_t& handle, uint64_t seed) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); IdxT nblks; IdxT nthreads; diff --git a/cpp/src/explainer/permutation_shap.cu b/cpp/src/explainer/permutation_shap.cu index 76ea8fa4e7..01fe01af1a 100644 --- a/cpp/src/explainer/permutation_shap.cu +++ b/cpp/src/explainer/permutation_shap.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -86,7 +86,7 @@ void permutation_shap_dataset_impl(const raft::handle_t& handle, bool row_major) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); // we calculate the number of rows in the dataset and then multiply by 2 since // we are adding a forward and backward permutation (see docstring in header file) @@ -127,7 +127,7 @@ void shap_main_effect_dataset_impl(const raft::handle_t& handle, bool row_major) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); // we calculate the number of elements in the dataset IdxT total_num_elements = (nrows_bg * ncols + nrows_bg) * ncols; @@ -180,7 +180,7 @@ void update_perm_shap_values_impl(const raft::handle_t& handle, const IdxT* idx) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); constexpr IdxT nthreads = 512; diff --git a/cpp/src/genetic/fitness.cuh b/cpp/src/genetic/fitness.cuh index 53bca068be..3049736116 100644 --- a/cpp/src/genetic/fitness.cuh +++ b/cpp/src/genetic/fitness.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -45,7 +45,7 @@ void weightedPearson(const raft::handle_t& h, { // Find Pearson's correlation coefficient - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector corr(n_samples * n_progs, stream); @@ -172,7 +172,7 @@ void weightedSpearman(const raft::handle_t& h, const math_t* W, math_t* out) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); // Get ranks for Y thrust::device_vector Ycopy(Y, Y + n_samples); @@ -235,7 +235,7 @@ void meanAbsoluteError(const raft::handle_t& h, const math_t* W, math_t* out) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector error(n_samples * n_progs, stream); rmm::device_scalar dWS(stream); math_t N = (math_t)n_samples; @@ -268,7 +268,7 @@ void meanSquareError(const raft::handle_t& h, const math_t* W, math_t* out) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector error(n_samples * n_progs, stream); rmm::device_scalar dWS(stream); math_t N = (math_t)n_samples; @@ -303,7 +303,7 @@ void rootMeanSquareError(const raft::handle_t& h, const math_t* W, math_t* out) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); // Find MSE meanSquareError(h, n_samples, n_progs, Y, Y_pred, W, out); @@ -321,7 +321,7 @@ void logLoss(const raft::handle_t& h, const math_t* W, math_t* out) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); // Logistic error per sample rmm::device_uvector error(n_samples * n_progs, stream); rmm::device_scalar dWS(stream); diff --git a/cpp/src/genetic/genetic.cu b/cpp/src/genetic/genetic.cu index ebb4c692d6..c3a1c677c1 100644 --- a/cpp/src/genetic/genetic.cu +++ b/cpp/src/genetic/genetic.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -121,7 +121,7 @@ void parallel_evolve(const raft::handle_t& h, const int generation, const int seed) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); auto n_progs = params.population_size; auto tour_size = params.tournament_size; auto n_tours = n_progs; // at least num_progs tournaments @@ -362,7 +362,7 @@ void symFit(const raft::handle_t& handle, program_t& final_progs, std::vector>& history) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Update arity map in params - Need to do this only here, as all operations will call Fit at // least once @@ -503,7 +503,7 @@ void symClfPredictProbs(const raft::handle_t& handle, const program_t& best_prog, float* output) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Assume output is of shape [n_rows, 2] in colMajor format execute(handle, best_prog, n_rows, 1, input, output); @@ -531,7 +531,7 @@ void symClfPredict(const raft::handle_t& handle, const program_t& best_prog, float* output) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Memory for probabilities rmm::device_uvector probs(2 * n_rows, stream); @@ -556,7 +556,7 @@ void symTransform(const raft::handle_t& handle, const int n_cols, float* output) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Execute final_progs(ordered by fitness) on input // output of size [n_rows,hall_of_fame] execute(handle, final_progs, n_rows, params.n_components, input, output); diff --git a/cpp/src/genetic/program.cu b/cpp/src/genetic/program.cu index 58be2710b4..abb18b9e3b 100644 --- a/cpp/src/genetic/program.cu +++ b/cpp/src/genetic/program.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -140,7 +140,7 @@ void execute(const raft::handle_t& h, const float* data, float* y_pred) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); dim3 blks(raft::ceildiv(n_rows, GENE_TPB), n_progs, 1); execute_kernel<<>>(d_progs, data, y_pred, (uint64_t)n_rows); @@ -156,7 +156,7 @@ void find_fitness(const raft::handle_t& h, const float* y, const float* sample_weights) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); // Compute predicted values rmm::device_uvector y_pred(n_rows, stream); @@ -176,7 +176,7 @@ void find_batched_fitness(const raft::handle_t& h, const float* y, const float* sample_weights) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector y_pred((uint64_t)n_rows * (uint64_t)n_progs, stream); execute(h, d_progs, n_rows, n_progs, data, y_pred.data()); @@ -194,7 +194,7 @@ void set_fitness(const raft::handle_t& h, const float* y, const float* sample_weights) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector score(1, stream); @@ -216,7 +216,7 @@ void set_batched_fitness(const raft::handle_t& h, const float* y, const float* sample_weights) { - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector score(n_progs, stream); diff --git a/cpp/src/glm/ols.cuh b/cpp/src/glm/ols.cuh index 0cefb7d9fd..45daa0cf3b 100644 --- a/cpp/src/glm/ols.cuh +++ b/cpp/src/glm/ols.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -54,7 +54,7 @@ void olsFit(const raft::handle_t& handle, int algo = 0, math_t* sample_weight = nullptr) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); auto cublas_handle = handle.get_cublas_handle(); auto cusolver_handle = handle.get_cusolver_dn_handle(); @@ -164,7 +164,7 @@ void gemmPredict(const raft::handle_t& handle, ASSERT(n_cols > 0, "gemmPredict: number of columns cannot be less than one"); ASSERT(n_rows > 0, "gemmPredict: number of rows cannot be less than one"); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); math_t alpha = math_t(1); math_t beta = math_t(0); raft::linalg::gemm(handle, diff --git a/cpp/src/glm/preprocess.cuh b/cpp/src/glm/preprocess.cuh index 9029a4719c..d77d8dd1ec 100644 --- a/cpp/src/glm/preprocess.cuh +++ b/cpp/src/glm/preprocess.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -47,7 +47,7 @@ void preProcessData(const raft::handle_t& handle, bool fit_intercept, math_t* sample_weight = nullptr) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); raft::common::nvtx::range fun_scope("ML::GLM::preProcessData-%d-%d", n_rows, n_cols); ASSERT(n_cols > 0, "Parameter n_cols: number of columns cannot be less than one"); ASSERT(n_rows > 1, "Parameter n_rows: number of rows cannot be less than two"); @@ -82,7 +82,7 @@ void postProcessData(const raft::handle_t& handle, math_t* mu_labels, bool fit_intercept) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); raft::common::nvtx::range fun_scope("ML::GLM::postProcessData-%d-%d", n_rows, n_cols); ASSERT(n_cols > 0, "Parameter n_cols: number of columns cannot be less than one"); ASSERT(n_rows > 1, "Parameter n_rows: number of rows cannot be less than two"); diff --git a/cpp/src/glm/qn/glm_base.cuh b/cpp/src/glm/qn/glm_base.cuh index 7db5685d68..20d5138d38 100644 --- a/cpp/src/glm/qn/glm_base.cuh +++ b/cpp/src/glm/qn/glm_base.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -31,7 +31,7 @@ inline void linearFwd(const raft::handle_t& handle, const SimpleMat& X, const SimpleDenseMat& W) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Forward pass: compute Z <- W * X.T + bias const bool has_bias = X.n != W.n; const int D = X.n; @@ -60,7 +60,7 @@ inline void linearBwd(const raft::handle_t& handle, const SimpleDenseMat& dZ, bool setZero) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Backward pass: // - compute G <- dZ * X.T // - for bias: Gb = mean(dZ, 1) diff --git a/cpp/src/glm/qn/mg/glm_base_mg.cuh b/cpp/src/glm/qn/mg/glm_base_mg.cuh index faf28a4eed..9d9d65ab8c 100644 --- a/cpp/src/glm/qn/mg/glm_base_mg.cuh +++ b/cpp/src/glm/qn/mg/glm_base_mg.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2023-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -32,7 +32,7 @@ inline void linearBwdMG(const raft::handle_t& handle, const int64_t n_samples, const int n_ranks) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Backward pass: // - compute G <- dZ * X.T // - for bias: Gb = mean(dZ, 1) diff --git a/cpp/src/glm/qn/mg/qn_mg.cuh b/cpp/src/glm/qn/mg/qn_mg.cuh index d151f1b1cc..3f512d3363 100644 --- a/cpp/src/glm/qn/mg/qn_mg.cuh +++ b/cpp/src/glm/qn/mg/qn_mg.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2023-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -37,7 +37,7 @@ int qn_fit_mg(const raft::handle_t& handle, int n_ranks, const Standardizer* stder_p = NULL) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); LBFGSParam opt_param(pams); SimpleVec w0(w0_data, loss.n_param); @@ -92,7 +92,7 @@ inline void qn_fit_x_mg(const raft::handle_t& handle, Dimensionality of w0 depends on loss, so we initialize it later. */ - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int N = X.m; int D = X.n; int n_targets = ML::GLM::detail::qn_is_classification(pams.loss) && C == 2 ? 1 : C; diff --git a/cpp/src/glm/qn/mg/standardization.cuh b/cpp/src/glm/qn/mg/standardization.cuh index 037e23a0ad..87d30268bd 100644 --- a/cpp/src/glm/qn/mg/standardization.cuh +++ b/cpp/src/glm/qn/mg/standardization.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -53,7 +53,7 @@ void vars(const raft::handle_t& handle, int D = X.n; int num_rows = X.m; bool col_major = (X.ord == COL_MAJOR); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto& comm = handle.get_comms(); rmm::device_uvector zero(D, handle.get_stream()); @@ -98,7 +98,7 @@ void mean_stddev(const raft::handle_t& handle, int D = X.n; int num_rows = X.m; bool col_major = (X.ord == COL_MAJOR); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto& comm = handle.get_comms(); if (col_major) { @@ -112,7 +112,7 @@ void mean_stddev(const raft::handle_t& handle, comm.sync_stream(stream); vars(handle, X, n_samples, mean_vector, stddev_vector); - raft::linalg::sqrt(stddev_vector, stddev_vector, D, handle.get_stream()); + raft::linalg::sqrt(stddev_vector, stddev_vector, D, handle.get_stream().get()); } template @@ -125,7 +125,7 @@ SimpleSparseMat get_sub_mat(const raft::handle_t& handle, end = end <= mat.m ? end : mat.m; int n_rows = end - start; int n_cols = mat.n; - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); RAFT_EXPECTS(start < end, "start index must be smaller than end index"); RAFT_EXPECTS(buff_row_ids.size() >= n_rows + 1, @@ -156,7 +156,7 @@ void mean(const raft::handle_t& handle, { int D = X.n; int num_rows = X.m; - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto& comm = handle.get_comms(); if (X.nnz == 0) { @@ -205,7 +205,7 @@ void mean_stddev(const raft::handle_t& handle, T* mean_vector, T* stddev_vector) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int D = X.n; mean(handle, X, n_samples, mean_vector); @@ -239,7 +239,7 @@ void mean_stddev(const raft::handle_t& handle, }; raft::linalg::binaryOp(stddev_vector, stddev_vector, mean_vector, X.n, submean_no_neg_op, stream); - raft::linalg::sqrt(stddev_vector, stddev_vector, X.n, handle.get_stream()); + raft::linalg::sqrt(stddev_vector, stddev_vector, X.n, handle.get_stream().get()); } struct inverse_op { @@ -265,7 +265,7 @@ struct Standardizer { int D = X.n; ASSERT(mean_std_buff.size() == 4 * D, "mean_std_buff size must be four times the dimension"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); mean.reset(mean_std_buff.data(), D); std.reset(mean_std_buff.data() + D, D); @@ -291,7 +291,7 @@ struct Standardizer { ASSERT(mean_std_buff.size() == 4 * vec_size, "mean_std_buff size must be four times the aligned size"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); T* p_ws = mean_std_buff.data(); @@ -330,14 +330,14 @@ struct Standardizer { Wweights.n, Wweights.m, mul_lambda, - handle.get_stream()); + handle.get_stream().get()); if (has_bias) { SimpleVec Wbias; col_ref(W, Wbias, D); - Wbias.assign_gemv(handle, -1, Wweights, false, mean, 1, handle.get_stream()); + Wbias.assign_gemv(handle, -1, Wweights, false, mean, 1, handle.get_stream().get()); } } @@ -346,7 +346,7 @@ struct Standardizer { const SimpleDenseMat& dZ, bool has_bias) const { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int D = mean.len; int n_targets = dZ.m; auto& comm = handle.get_comms(); diff --git a/cpp/src/glm/qn/qn.cuh b/cpp/src/glm/qn/qn.cuh index 5353411d34..d26327ac52 100644 --- a/cpp/src/glm/qn/qn.cuh +++ b/cpp/src/glm/qn/qn.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -36,7 +36,7 @@ int qn_fit(const raft::handle_t& handle, T* fx, int* num_iters) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); LBFGSParam opt_param(pams); SimpleVec w0(w0_data, loss.n_param); @@ -100,7 +100,7 @@ inline void qn_fit_x(const raft::handle_t& handle, Dimensionality of w0 depends on loss, so we initialize it later. */ - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int N = X.m; int D = X.n; int n_targets = qn_is_classification(pams.loss) && C == 2 ? 1 : C; @@ -253,7 +253,7 @@ template void qn_predict( const raft::handle_t& handle, const qn_params& pams, SimpleMat& X, int C, T* params, T* preds) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); bool is_class = qn_is_classification(pams.loss); int n_targets = is_class && C == 2 ? 1 : C; rmm::device_uvector scores(checked_mul(n_targets, X.m), stream); diff --git a/cpp/src/glm/qn/qn_solvers.cuh b/cpp/src/glm/qn/qn_solvers.cuh index 770b8edef1..17c07f4170 100644 --- a/cpp/src/glm/qn/qn_solvers.cuh +++ b/cpp/src/glm/qn/qn_solvers.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -413,7 +413,7 @@ inline int qn_minimize(const raft::handle_t& handle, const rapids_logger::level_enum verbosity = 0) { // TODO should the worksapce allocation happen outside? - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); OPT_RETCODE ret; if (l1 == 0.0) { rmm::device_uvector tmp(lbfgs_workspace_size(opt_param, x.len), stream); diff --git a/cpp/src/glm/qn/simple_mat/dense.hpp b/cpp/src/glm/qn/simple_mat/dense.hpp index f08036a8de..a257b7290a 100644 --- a/cpp/src/glm/qn/simple_mat/dense.hpp +++ b/cpp/src/glm/qn/simple_mat/dense.hpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -12,6 +12,8 @@ #include #include +#include + #include #include // #TODO: Replace with public header when ready @@ -340,7 +342,7 @@ std::ostream& operator<<(std::ostream& os, const SimpleDenseMat& mat) { os << "ord=" << (mat.ord == COL_MAJOR ? "CM" : "RM") << "\n"; std::vector out(mat.len); - raft::update_host(&out[0], mat.data, mat.len, rmm::cuda_stream_default); + raft::update_host(&out[0], mat.data, mat.len, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); raft::interruptible::synchronize(rmm::cuda_stream_view()); if (mat.ord == COL_MAJOR) { for (int r = 0; r < mat.m; r++) { diff --git a/cpp/src/glm/qn/simple_mat/sparse.hpp b/cpp/src/glm/qn/simple_mat/sparse.hpp index 2b0c6ae60f..d5f8ef79b6 100644 --- a/cpp/src/glm/qn/simple_mat/sparse.hpp +++ b/cpp/src/glm/qn/simple_mat/sparse.hpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -18,6 +18,8 @@ #include +#include + #include #include @@ -176,9 +178,11 @@ std::ostream& operator<<(std::ostream& os, const SimpleSparseMat& mat) std::vector values(mat.nnz); std::vector cols(mat.nnz); std::vector row_ids(mat.m + 1); - raft::update_host(&values[0], mat.values, mat.nnz, rmm::cuda_stream_default); - raft::update_host(&cols[0], mat.cols, mat.nnz, rmm::cuda_stream_default); - raft::update_host(&row_ids[0], mat.row_ids, mat.m + 1, rmm::cuda_stream_default); + raft::update_host( + &values[0], mat.values, mat.nnz, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); + raft::update_host(&cols[0], mat.cols, mat.nnz, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); + raft::update_host( + &row_ids[0], mat.row_ids, mat.m + 1, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); raft::interruptible::synchronize(rmm::cuda_stream_view()); int i, row_end = 0; diff --git a/cpp/src/glm/qn_mg.cu b/cpp/src/glm/qn_mg.cu index c6ac17a5c5..142213ecdb 100644 --- a/cpp/src/glm/qn_mg.cu +++ b/cpp/src/glm/qn_mg.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2023-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -43,7 +43,7 @@ namespace opg { template std::vector distinct_mg(const raft::handle_t& handle, T* y, size_t n) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); raft::comms::comms_t const& comm = raft::resource::get_comms(handle); int rank = comm.get_rank(); int n_ranks = comm.get_size(); diff --git a/cpp/src/glm/ridge.cuh b/cpp/src/glm/ridge.cuh index 3d72158472..4f4574eff6 100644 --- a/cpp/src/glm/ridge.cuh +++ b/cpp/src/glm/ridge.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -41,7 +41,7 @@ void ridgeSolve(const raft::handle_t& handle, int n_alpha, math_t* w) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto cublasH = handle.get_cublas_handle(); auto cusolverH = handle.get_cusolver_dn_handle(); @@ -87,7 +87,7 @@ void ridgeSVD(const raft::handle_t& handle, int n_alpha, math_t* w) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto cublasH = handle.get_cublas_handle(); auto cusolverH = handle.get_cusolver_dn_handle(); @@ -116,7 +116,7 @@ void ridgeEig(const raft::handle_t& handle, int n_alpha, math_t* w) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto cublasH = handle.get_cublas_handle(); auto cusolverH = handle.get_cusolver_dn_handle(); @@ -165,7 +165,7 @@ void ridgeFit(const raft::handle_t& handle, int algo = 0, math_t* sample_weight = nullptr) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); auto cublas_handle = handle.get_cublas_handle(); auto cusolver_handle = handle.get_cusolver_dn_handle(); diff --git a/cpp/src/hdbscan/condensed_hierarchy.cu b/cpp/src/hdbscan/condensed_hierarchy.cu index 36dae7aff2..105f94591c 100644 --- a/cpp/src/hdbscan/condensed_hierarchy.cu +++ b/cpp/src/hdbscan/condensed_hierarchy.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -84,7 +84,7 @@ CondensedHierarchy::CondensedHierarchy(const raft::handle_t& auto parents_ptr = thrust::device_pointer_cast(parents.data()); auto parents_min_max = thrust::minmax_element( - thrust::cuda::par.on(handle.get_stream()), parents_ptr, parents_ptr + n_edges); + thrust::cuda::par.on(handle.get_stream().get()), parents_ptr, parents_ptr + n_edges); auto min_cluster = *parents_min_max.first; auto max_cluster = *parents_min_max.second; @@ -94,7 +94,7 @@ CondensedHierarchy::CondensedHierarchy(const raft::handle_t& cuda::std::make_tuple(parents.begin(), children.begin(), sizes.begin())); auto sort_values = thrust::make_zip_iterator(cuda::std::make_tuple(lambdas.begin())); - thrust::sort_by_key(thrust::cuda::par.on(handle.get_stream()), + thrust::sort_by_key(thrust::cuda::par.on(handle.get_stream().get()), sort_keys, sort_keys + n_edges, sort_values, @@ -137,7 +137,7 @@ void CondensedHierarchy::condense(value_idx* full_parents, value_idx* full_sizes, value_idx size) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); if (size == -1) size = 4 * (n_leaves - 1) + 2; diff --git a/cpp/src/hdbscan/detail/condense.cuh b/cpp/src/hdbscan/detail/condense.cuh index 8f4148d382..b1bc253fd3 100644 --- a/cpp/src/hdbscan/detail/condense.cuh +++ b/cpp/src/hdbscan/detail/condense.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -100,7 +100,7 @@ void _build_condensed_hierarchy(const raft::handle_t& handle, rmm::device_uvector& out_lambda, rmm::device_uvector& out_size) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); value_idx root = 2 * (n_leaves - 1); value_idx n_samples = n_leaves; value_idx next_label = n_samples + 1; @@ -243,7 +243,7 @@ void build_condensed_hierarchy(const raft::handle_t& handle, int n_leaves, Common::CondensedHierarchy& condensed_tree) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); // Root is the last edge in the dendrogram diff --git a/cpp/src/hdbscan/detail/membership.cuh b/cpp/src/hdbscan/detail/membership.cuh index 3272ee6e14..8bde9ce1c6 100644 --- a/cpp/src/hdbscan/detail/membership.cuh +++ b/cpp/src/hdbscan/detail/membership.cuh @@ -42,7 +42,7 @@ void get_probabilities(const raft::handle_t& handle, const value_idx* labels, value_t* probabilities) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); auto parents = condensed_tree.get_parents(); diff --git a/cpp/src/hdbscan/detail/predict.cuh b/cpp/src/hdbscan/detail/predict.cuh index 68cdbff80c..2719ce0a4f 100644 --- a/cpp/src/hdbscan/detail/predict.cuh +++ b/cpp/src/hdbscan/detail/predict.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -52,7 +52,7 @@ void _find_neighbor_and_lambda(const raft::handle_t& handle, value_idx* min_mr_inds, value_t* prediction_lambdas) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); // Buffer for storing the minimum mutual reachability distances @@ -111,7 +111,7 @@ void _find_cluster_and_probability(const raft::handle_t& handle, value_idx* out_labels, value_t* out_probabilities) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); auto parents = condensed_tree.get_parents(); @@ -154,7 +154,7 @@ void _compute_knn_and_nearest_neighbor(const raft::handle_t& handle, value_t* prediction_lambdas, ML::distance::DistanceType metric) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); size_t m = prediction_data.n_rows; size_t n = prediction_data.n_cols; value_t* input_core_dists = prediction_data.get_core_dists(); diff --git a/cpp/src/hdbscan/detail/reachability.cuh b/cpp/src/hdbscan/detail/reachability.cuh index c1bdf2e624..5e1d45c799 100644 --- a/cpp/src/hdbscan/detail/reachability.cuh +++ b/cpp/src/hdbscan/detail/reachability.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -132,7 +132,7 @@ void _compute_core_dists(const raft::handle_t& handle, RAFT_EXPECTS(metric == ML::distance::DistanceType::L2SqrtExpanded, "Currently only L2 expanded distance is supported"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_uvector inds(min_samples * m, stream); rmm::device_uvector dists(min_samples * m, stream); diff --git a/cpp/src/hdbscan/detail/select.cuh b/cpp/src/hdbscan/detail/select.cuh index 4f6627e05b..643936119c 100644 --- a/cpp/src/hdbscan/detail/select.cuh +++ b/cpp/src/hdbscan/detail/select.cuh @@ -63,7 +63,7 @@ void perform_bfs(const raft::handle_t& handle, int n_clusters, Bfs_Kernel bfs_kernel) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto thrust_policy = handle.get_thrust_policy(); rmm::device_uvector next_frontier(n_clusters, stream); @@ -105,7 +105,7 @@ void parent_csr(const raft::handle_t& handle, Common::CondensedHierarchy& cluster_tree, value_idx* indptr) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto parents = cluster_tree.get_parents(); auto children = cluster_tree.get_children(); @@ -307,7 +307,7 @@ void cluster_epsilon_search(const raft::handle_t& handle, const bool allow_single_cluster, const int n_selected_clusters) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto thrust_policy = handle.get_thrust_policy(); auto parents = cluster_tree.get_parents(); auto children = cluster_tree.get_children(); diff --git a/cpp/src/hdbscan/detail/soft_clustering.cuh b/cpp/src/hdbscan/detail/soft_clustering.cuh index 75f1165142..65bfd217c7 100644 --- a/cpp/src/hdbscan/detail/soft_clustering.cuh +++ b/cpp/src/hdbscan/detail/soft_clustering.cuh @@ -59,7 +59,7 @@ void dist_membership_vector(const raft::handle_t& handle, size_t batch_size, bool softmax = false) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); rmm::device_uvector exemplars_dense(n_exemplars * n, stream); @@ -161,7 +161,7 @@ void all_points_outlier_membership_vector( value_t* outlier_membership_vec, bool softmax) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); auto parents = condensed_tree.get_parents(); @@ -266,7 +266,7 @@ void outlier_membership_vector(const raft::handle_t& handle, value_t* outlier_membership_vec, bool softmax) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); auto parents = condensed_tree.get_parents(); @@ -391,7 +391,7 @@ void all_points_membership_vectors(const raft::handle_t& handle, value_t* membership_vec, size_t batch_size) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); size_t m = prediction_data.n_rows; @@ -513,7 +513,7 @@ void membership_vector(const raft::handle_t& handle, RAFT_EXPECTS(metric == ML::distance::DistanceType::L2SqrtExpanded, "Currently only L2 expanded distance is supported"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); size_t m = prediction_data.n_rows; diff --git a/cpp/src/hdbscan/detail/stabilities.cuh b/cpp/src/hdbscan/detail/stabilities.cuh index 814f4cac35..8d7d6523bd 100644 --- a/cpp/src/hdbscan/detail/stabilities.cuh +++ b/cpp/src/hdbscan/detail/stabilities.cuh @@ -59,7 +59,7 @@ void compute_stabilities(const raft::handle_t& handle, auto n_clusters = condensed_tree.get_n_clusters(); auto n_leaves = condensed_tree.get_n_leaves(); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); rmm::device_uvector sorted_parents(n_edges, stream); diff --git a/cpp/src/hdbscan/detail/utils.cuh b/cpp/src/hdbscan/detail/utils.cuh index 6aff7389fe..b4d74c5c39 100644 --- a/cpp/src/hdbscan/detail/utils.cuh +++ b/cpp/src/hdbscan/detail/utils.cuh @@ -157,7 +157,7 @@ void parent_csr(const raft::handle_t& handle, value_idx* sorted_parents, value_idx* indptr) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto thrust_policy = handle.get_thrust_policy(); auto children = condensed_tree.get_children(); diff --git a/cpp/src/hdbscan/prediction_data.cu b/cpp/src/hdbscan/prediction_data.cu index 9490873d65..c2824afe36 100644 --- a/cpp/src/hdbscan/prediction_data.cu +++ b/cpp/src/hdbscan/prediction_data.cu @@ -96,7 +96,7 @@ void generate_prediction_data(const raft::handle_t& handle, int n_selected_clusters, PredictionData& prediction_data) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto exec_policy = handle.get_thrust_policy(); auto counting = thrust::make_counting_iterator(0); diff --git a/cpp/src/holtwinters/internal/hw_decompose.cuh b/cpp/src/holtwinters/internal/hw_decompose.cuh index 5254b0efe7..1c9b46f617 100644 --- a/cpp/src/holtwinters/internal/hw_decompose.cuh +++ b/cpp/src/holtwinters/internal/hw_decompose.cuh @@ -55,7 +55,7 @@ void conv1d(const raft::handle_t& handle, <<>>(input, batch_size, filter, filter_size, output, output_size); + handle.get_stream().get()>>>(input, batch_size, filter, filter_size, output, output_size); } // https://github.com/NVIDIA/cuml/issues/891 @@ -104,7 +104,7 @@ void season_mean(const raft::handle_t& handle, int half_filter_size, ML::SeasonalType seasonal) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); bool is_additive = seasonal == ML::SeasonalType::ADDITIVE; season_mean_kernel <<>>( @@ -151,7 +151,7 @@ void batched_ls(const raft::handle_t& handle, Dtype* level, Dtype* trend) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); cublasHandle_t cublas_h = handle.get_cublas_handle(); cusolverDnHandle_t cusolver_h = handle.get_cusolver_dn_handle(); @@ -252,7 +252,7 @@ void stl_decomposition_gpu(const raft::handle_t& handle, Dtype* start_season, ML::SeasonalType seasonal) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); cublasHandle_t cublas_h = handle.get_cublas_handle(); const int end = ML::checked_mul(start_periods, frequency); diff --git a/cpp/src/holtwinters/internal/hw_eval.cuh b/cpp/src/holtwinters/internal/hw_eval.cuh index aaa2bb669d..eac1ae6b21 100644 --- a/cpp/src/holtwinters/internal/hw_eval.cuh +++ b/cpp/src/holtwinters/internal/hw_eval.cuh @@ -237,7 +237,7 @@ void holtwinters_eval_gpu(const raft::handle_t& handle, Dtype* error, ML::SeasonalType seasonal) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int total_blocks = GET_NUM_BLOCKS(batch_size); int threads_per_block = GET_THREADS_PER_BLOCK(batch_size); diff --git a/cpp/src/holtwinters/internal/hw_forecast.cuh b/cpp/src/holtwinters/internal/hw_forecast.cuh index 14069cec72..5f67a4de1c 100644 --- a/cpp/src/holtwinters/internal/hw_forecast.cuh +++ b/cpp/src/holtwinters/internal/hw_forecast.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -68,7 +68,7 @@ void holtwinters_forecast_gpu(const raft::handle_t& handle, const Dtype* season_coef, ML::SeasonalType seasonal) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int total_blocks = GET_NUM_BLOCKS(batch_size); int threads_per_block = GET_THREADS_PER_BLOCK(batch_size); diff --git a/cpp/src/holtwinters/internal/hw_optim.cuh b/cpp/src/holtwinters/internal/hw_optim.cuh index 49ba0feb36..e356816910 100644 --- a/cpp/src/holtwinters/internal/hw_optim.cuh +++ b/cpp/src/holtwinters/internal/hw_optim.cuh @@ -855,7 +855,7 @@ void holtwinters_optim_gpu(const raft::handle_t& handle, ML::SeasonalType seasonal, const ML::OptimParams optim_params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // int total_blocks = GET_NUM_BLOCKS(batch_size); // int threads_per_block = GET_THREADS_PER_BLOCK(batch_size); diff --git a/cpp/src/holtwinters/runner.cuh b/cpp/src/holtwinters/runner.cuh index f4235175cc..a8d278e4c8 100644 --- a/cpp/src/holtwinters/runner.cuh +++ b/cpp/src/holtwinters/runner.cuh @@ -31,7 +31,7 @@ void HWTranspose(const raft::handle_t& handle, Dtype* data_in, int m, int n, Dty ASSERT(!(!data_in || !data_out || n < 1 || m < 1), "HW error in in line %d", __LINE__); const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); cublasHandle_t cublas_h = handle_impl.get_cublas_handle(); raft::linalg::transpose(handle, data_in, data_out, n, m, stream); @@ -110,7 +110,7 @@ void HoltWintersDecompose(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); cublasHandle_t cublas_h = handle_impl.get_cublas_handle(); if (start_level != nullptr && start_trend == nullptr && @@ -160,7 +160,7 @@ void HoltWintersEval(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); ASSERT(!((!start_trend) != (!beta) || (!start_season) != (!gamma)), "HW error in in line %d", @@ -219,7 +219,7 @@ void HoltWintersOptim(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); // default values OptimParams optim_params_; @@ -298,7 +298,7 @@ void HoltWintersForecast(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); ASSERT(!(!level_coef && !trend_coef && !season_coef), "HW error in in line %d", __LINE__); ASSERT(!(season_coef && frequency < 2), "HW error in in line %d", __LINE__); @@ -324,7 +324,7 @@ void HoltWintersFitHelper(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); bool optim_alpha = true, optim_beta = true, optim_gamma = true; // initial values for alpha, beta and gamma @@ -425,7 +425,7 @@ void HoltWintersForecastHelper(const raft::handle_t& handle, { const raft::handle_t& handle_impl = handle; raft::stream_syncer _(handle_impl); - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); bool optim_beta = true, optim_gamma = true; diff --git a/cpp/src/isolation_forest/isolation_forest.cu b/cpp/src/isolation_forest/isolation_forest.cu index 08a1d0b560..dbb8e56c5e 100644 --- a/cpp/src/isolation_forest/isolation_forest.cu +++ b/cpp/src/isolation_forest/isolation_forest.cu @@ -219,7 +219,7 @@ void predict(const raft::handle_t& handle, rapids_logger::level_enum verbosity) { ML::default_logger().set_level(verbosity); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // First compute anomaly scores rmm::device_uvector scores(n_rows, stream); @@ -245,7 +245,7 @@ void predict(const raft::handle_t& handle, rapids_logger::level_enum verbosity) { ML::default_logger().set_level(verbosity); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // First compute anomaly scores rmm::device_uvector scores(n_rows, stream); diff --git a/cpp/src/isolation_forest/isolation_forest.cuh b/cpp/src/isolation_forest/isolation_forest.cuh index a25299a5e6..899d89fbda 100644 --- a/cpp/src/isolation_forest/isolation_forest.cuh +++ b/cpp/src/isolation_forest/isolation_forest.cuh @@ -140,7 +140,7 @@ class IsolationForest { T* avg_path_lengths) const { raft::common::nvtx::range fun_scope("IF::compute_path_lengths @isolation_forest.cuh"); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int threads = 256; size_t blocks = (n_rows + threads - 1) / threads; @@ -161,7 +161,7 @@ class IsolationForest { size_t n_rows, T* scores) const { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); T c_n = static_cast(model->c_normalization); if (c_n <= T(0)) { diff --git a/cpp/src/isolation_forest/isolation_tree_builder.cuh b/cpp/src/isolation_forest/isolation_tree_builder.cuh index d9f6585d9d..2da97b467f 100644 --- a/cpp/src/isolation_forest/isolation_tree_builder.cuh +++ b/cpp/src/isolation_forest/isolation_tree_builder.cuh @@ -394,7 +394,7 @@ void build_isolation_forest_global(const raft::handle_t& handle, int* tree_n_nodes, int* tree_max_depth) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); size_t subsample_buffer_size = static_cast(n_trees) * max_samples * max_features; rmm::device_uvector subsample_buffer(subsample_buffer_size, stream); @@ -455,7 +455,7 @@ void compact_global_isolation_forest(const raft::handle_t& handle, std::vector& h_tree_n_nodes, std::vector& h_tree_max_depth) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); h_tree_n_nodes.resize(n_trees); h_tree_max_depth.resize(n_trees); diff --git a/cpp/src/knn/knn.cu b/cpp/src/knn/knn.cu index d2e1021461..be0a4dd15b 100644 --- a/cpp/src/knn/knn.cu +++ b/cpp/src/knn/knn.cu @@ -247,7 +247,7 @@ void approx_knn_build_index(raft::handle_t& handle, auto ivf_ft_pams = dynamic_cast(params); auto ivf_pq_pams = dynamic_cast(params); - auto stream = raft::resource::get_cuda_stream(handle); + auto stream = raft::resource::get_cuda_stream(handle).get(); // For correlation: preprocess (center + normalize), use InnerProduct, then revert if (metric == ML::distance::DistanceType::CorrelationExpanded) { @@ -319,7 +319,7 @@ void approx_knn_search(raft::handle_t& handle, float* query_array, int n) { - auto stream = raft::resource::get_cuda_stream(handle); + auto stream = raft::resource::get_cuda_stream(handle).get(); // Get dimension from index int D = index->pimpl->ivf_flat ? index->pimpl->ivf_flat->dim() : index->pimpl->ivf_pq->dim(); @@ -386,7 +386,7 @@ void approx_knn_search(raft::handle_t& handle, distances, n * k, raft::pow_const_op(p), - raft::resource::get_cuda_stream(handle)); + raft::resource::get_cuda_stream(handle).get()); } // Post-process correlation: convert inner product to correlation distance @@ -405,7 +405,7 @@ void knn_classify(raft::handle_t& handle, int k, float* sample_weight) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); std::vector> uniq_labels_v; std::vector uniq_labels(y.size()); @@ -464,7 +464,7 @@ void knn_class_proba(raft::handle_t& handle, int k, float* sample_weight) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); std::vector> uniq_labels_v; std::vector uniq_labels(y.size()); diff --git a/cpp/src/knn/knn_opg_common.cuh b/cpp/src/knn/knn_opg_common.cuh index 954516124c..8f3b1de9f0 100644 --- a/cpp/src/knn/knn_opg_common.cuh +++ b/cpp/src/knn/knn_opg_common.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -373,7 +373,7 @@ void broadcast_query(opg_knn_work& work, in_t* broadcast, size_t broadcast_size) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); if (part_rank == work.my_rank) { // Sender: send to all other idx ranks for (int rank : work.idxRanks) { @@ -506,13 +506,13 @@ void copy_label_outputs_from_index_parts(opg_knn_param - <<>>(work.res.data() + (o * n_labels), - work.res_I.data(), - parts_d.data(), - offsets_d.data(), - batch_size, - n_parts, - n_labels); + <<>>(work.res.data() + (o * n_labels), + work.res_I.data(), + parts_d.data(), + offsets_d.data(), + batch_size, + n_parts, + n_labels); } handle.sync_stream(handle.get_stream()); RAFT_CUDA_TRY(cudaPeekAtLastError()); @@ -536,7 +536,7 @@ void exchange_results(opg_knn_param& params, size_t batch_size) { size_t batch_elms = batch_size * params.k; - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); if (part_rank != work.my_rank) { // Sender: send local KNN results to part_rank handle.get_comms().device_send(work.res_I.data(), batch_elms, part_rank, stream); @@ -643,23 +643,23 @@ void reduce(opg_knn_param& params, size_t processed_in_part, size_t batch_size) { - rmm::device_uvector trans(work.idxRanks.size(), handle.get_stream()); - RAFT_CUDA_TRY( - cudaMemsetAsync(trans.data(), 0, work.idxRanks.size() * sizeof(trans_t), handle.get_stream())); + rmm::device_uvector trans(work.idxRanks.size(), handle.get_stream().get()); + RAFT_CUDA_TRY(cudaMemsetAsync( + trans.data(), 0, work.idxRanks.size() * sizeof(trans_t), handle.get_stream().get())); size_t batch_offset = processed_in_part * params.k; ind_t* indices = nullptr; dist_t* distances = nullptr; - rmm::device_uvector indices_b(0, handle.get_stream()); - rmm::device_uvector distances_b(0, handle.get_stream()); + rmm::device_uvector indices_b(0, handle.get_stream().get()); + rmm::device_uvector distances_b(0, handle.get_stream().get()); if (params.knn_op == knn_operation::knn) { indices = params.out_I->at(part_idx)->ptr + batch_offset; distances = params.out_D->at(part_idx)->ptr + batch_offset; } else { - indices_b.resize(batch_size * params.k, handle.get_stream()); + indices_b.resize(batch_size * params.k, handle.get_stream().get()); distances_b.resize(batch_size * params.k, handle.get_stream()); indices = indices_b.data(); distances = distances_b.data(); @@ -811,17 +811,18 @@ void merge_labels(opg_knn_param_t& params, raft::update_device( parts_to_ranks_d.data(), parts_to_ranks_h.data(), parts_to_ranks_h.size(), handle.get_stream()); - merge_labels_kernel<<>>(output, - knn_indices, - unmerged_outputs, - unmerged_knn_indices, - offsets_d.data(), - parts_to_ranks_d.data(), - params.k, - params.n_outputs, - n_labels, - work.idxPartsToRanks.size(), - work.idxRanks.size()); + merge_labels_kernel + <<>>(output, + knn_indices, + unmerged_outputs, + unmerged_knn_indices, + offsets_d.data(), + parts_to_ranks_d.data(), + params.k, + params.n_outputs, + n_labels, + work.idxPartsToRanks.size(), + work.idxRanks.size()); } /*! diff --git a/cpp/src/metrics/accuracy_score.cu b/cpp/src/metrics/accuracy_score.cu index 8c2371de2c..0acc1139ee 100644 --- a/cpp/src/metrics/accuracy_score.cu +++ b/cpp/src/metrics/accuracy_score.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -18,7 +18,7 @@ float accuracy_score_py(const raft::handle_t& handle, const int* ref_predictions, int n) { - return raft::stats::accuracy(predictions, ref_predictions, n, handle.get_stream()); + return raft::stats::accuracy(predictions, ref_predictions, n, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/adjusted_rand_index.cu b/cpp/src/metrics/adjusted_rand_index.cu index 3edbe7f0c9..c86b0ca094 100644 --- a/cpp/src/metrics/adjusted_rand_index.cu +++ b/cpp/src/metrics/adjusted_rand_index.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -18,7 +18,7 @@ double adjusted_rand_index(const raft::handle_t& handle, const int64_t n) { return raft::stats::adjusted_rand_index( - y, y_hat, n, handle.get_stream()); + y, y_hat, n, handle.get_stream().get()); } double adjusted_rand_index(const raft::handle_t& handle, @@ -27,7 +27,7 @@ double adjusted_rand_index(const raft::handle_t& handle, const int n) { return raft::stats::adjusted_rand_index( - y, y_hat, n, handle.get_stream()); + y, y_hat, n, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/completeness_score.cu b/cpp/src/metrics/completeness_score.cu index 38437e3806..d7905a9ed7 100644 --- a/cpp/src/metrics/completeness_score.cu +++ b/cpp/src/metrics/completeness_score.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -21,7 +21,7 @@ double completeness_score(const raft::handle_t& handle, const int upper_class_range) { return raft::stats::homogeneity_score( - y_hat, y, n, lower_class_range, upper_class_range, handle.get_stream()); + y_hat, y, n, lower_class_range, upper_class_range, handle.get_stream().get()); } } // namespace Metrics diff --git a/cpp/src/metrics/entropy.cu b/cpp/src/metrics/entropy.cu index 8bb41245dc..d13c6facb6 100644 --- a/cpp/src/metrics/entropy.cu +++ b/cpp/src/metrics/entropy.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -18,7 +18,8 @@ double entropy(const raft::handle_t& handle, const int lower_class_range, const int upper_class_range) { - return raft::stats::entropy(y, n, lower_class_range, upper_class_range, handle.get_stream()); + return raft::stats::entropy( + y, n, lower_class_range, upper_class_range, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/homogeneity_score.cu b/cpp/src/metrics/homogeneity_score.cu index d875863d8a..aaeb3d488e 100644 --- a/cpp/src/metrics/homogeneity_score.cu +++ b/cpp/src/metrics/homogeneity_score.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -21,7 +21,7 @@ double homogeneity_score(const raft::handle_t& handle, const int upper_class_range) { return raft::stats::homogeneity_score( - y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream()); + y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/kl_divergence.cu b/cpp/src/metrics/kl_divergence.cu index 2e6ad6f885..0632071535 100644 --- a/cpp/src/metrics/kl_divergence.cu +++ b/cpp/src/metrics/kl_divergence.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -15,12 +15,12 @@ namespace Metrics { double kl_divergence(const raft::handle_t& handle, const double* y, const double* y_hat, int n) { - return raft::stats::kl_divergence(y, y_hat, n, handle.get_stream()); + return raft::stats::kl_divergence(y, y_hat, n, handle.get_stream().get()); } float kl_divergence(const raft::handle_t& handle, const float* y, const float* y_hat, int n) { - return raft::stats::kl_divergence(y, y_hat, n, handle.get_stream()); + return raft::stats::kl_divergence(y, y_hat, n, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/mutual_info_score.cu b/cpp/src/metrics/mutual_info_score.cu index d52c2929c7..274c34eaf1 100644 --- a/cpp/src/metrics/mutual_info_score.cu +++ b/cpp/src/metrics/mutual_info_score.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -21,7 +21,7 @@ double mutual_info_score(const raft::handle_t& handle, const int upper_class_range) { return raft::stats::mutual_info_score( - y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream()); + y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream().get()); } } // namespace Metrics diff --git a/cpp/src/metrics/r2_score.cu b/cpp/src/metrics/r2_score.cu index 5eda722b6d..aecc273e6b 100644 --- a/cpp/src/metrics/r2_score.cu +++ b/cpp/src/metrics/r2_score.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,12 +14,12 @@ namespace Metrics { float r2_score_py(const raft::handle_t& handle, float* y, float* y_hat, int n) { - return raft::stats::r2_score(y, y_hat, n, handle.get_stream()); + return raft::stats::r2_score(y, y_hat, n, handle.get_stream().get()); } double r2_score_py(const raft::handle_t& handle, double* y, double* y_hat, int n) { - return raft::stats::r2_score(y, y_hat, n, handle.get_stream()); + return raft::stats::r2_score(y, y_hat, n, handle.get_stream().get()); } } // namespace Metrics diff --git a/cpp/src/metrics/rand_index.cu b/cpp/src/metrics/rand_index.cu index daf2675511..102a2bbf22 100644 --- a/cpp/src/metrics/rand_index.cu +++ b/cpp/src/metrics/rand_index.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -15,7 +15,7 @@ namespace Metrics { double rand_index(const raft::handle_t& handle, const double* y, const double* y_hat, int n) { - return raft::stats::rand_index(y, y_hat, (uint64_t)n, handle.get_stream()); + return raft::stats::rand_index(y, y_hat, (uint64_t)n, handle.get_stream().get()); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/metrics/v_measure.cu b/cpp/src/metrics/v_measure.cu index afd32c4b7b..d6881e1eaa 100644 --- a/cpp/src/metrics/v_measure.cu +++ b/cpp/src/metrics/v_measure.cu @@ -1,6 +1,6 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -22,7 +22,7 @@ double v_measure(const raft::handle_t& handle, double beta) { return raft::stats::v_measure( - y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream(), beta); + y, y_hat, n, lower_class_range, upper_class_range, handle.get_stream().get(), beta); } } // namespace Metrics } // namespace ML diff --git a/cpp/src/randomforest/randomforest.cuh b/cpp/src/randomforest/randomforest.cuh index b44c1d3c80..4bec399b0f 100644 --- a/cpp/src/randomforest/randomforest.cuh +++ b/cpp/src/randomforest/randomforest.cuh @@ -363,7 +363,7 @@ class RandomForest { int stream_id = omp_get_thread_num(); auto s = handle.get_stream_from_stream_pool(stream_id); - auto& selected_rows = row_sampler.sample(i, stream_id, s); + auto& selected_rows = row_sampler.sample(i, stream_id, s.get()); /* Build individual tree in the forest. - input is a pointer to orig data that have n_cols features and n_rows rows. @@ -375,7 +375,7 @@ class RandomForest { */ forest->trees[i] = DT::DecisionTree::fit(handle, - s, + s.get(), input, n_cols, n_rows, @@ -415,7 +415,7 @@ class RandomForest { ML::default_logger().set_level(verbosity); this->error_checking(input, predictions, n_rows, n_cols, true); std::vector h_predictions(n_rows); - cudaStream_t stream = user_handle.get_stream(); + cudaStream_t stream = user_handle.get_stream().get(); std::vector h_input(std::size_t(n_rows) * n_cols); raft::update_host(h_input.data(), input, std::size_t(n_rows) * n_cols, stream); @@ -480,7 +480,7 @@ class RandomForest { int rf_type = RF_type::CLASSIFICATION) { ML::default_logger().set_level(verbosity); - cudaStream_t stream = user_handle.get_stream(); + cudaStream_t stream = user_handle.get_stream().get(); RF_metrics stats; if (rf_type == RF_type::CLASSIFICATION) { // task classifiation: get classification metrics float accuracy = raft::stats::accuracy(predictions, ref_labels, n_rows, stream); diff --git a/cpp/src/solver/cd.cuh b/cpp/src/solver/cd.cuh index 1da8b4a764..1f1676229a 100644 --- a/cpp/src/solver/cd.cuh +++ b/cpp/src/solver/cd.cuh @@ -146,7 +146,7 @@ int cdFit(const raft::handle_t& handle, ASSERT(loss == ML::loss_funct::SQRD_LOSS, "Parameter loss: Only SQRT_LOSS function is supported for now"); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); rmm::device_uvector residual(n_rows, stream); rmm::device_uvector squared(n_cols, stream); rmm::device_uvector mu_input(0, stream); @@ -343,7 +343,8 @@ void cdPredict(const raft::handle_t& handle, ASSERT(loss == ML::loss_funct::SQRD_LOSS, "Parameter loss: Only SQRT_LOSS function is supported for now"); - Functions::linearRegH(handle, input, n_rows, n_cols, coef, preds, intercept, handle.get_stream()); + Functions::linearRegH( + handle, input, n_rows, n_cols, coef, preds, intercept, handle.get_stream().get()); } }; // namespace Solver diff --git a/cpp/src/solver/lars_impl.cuh b/cpp/src/solver/lars_impl.cuh index 5adfde77cf..cac8d8912e 100644 --- a/cpp/src/solver/lars_impl.cuh +++ b/cpp/src/solver/lars_impl.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -891,7 +891,7 @@ void larsFit(const raft::handle_t& handle, if (ld_X == 0) ld_X = n_rows; if (Gram && ld_G == 0) ld_G = n_cols; - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // We will use either U_buffer.data() to store the Cholesky factorization, or // store it in place at Gram. Pointer U will point to the actual storage. @@ -1092,7 +1092,7 @@ void larsPredict(const raft::handle_t& handle, math_t intercept, math_t* preds) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); rmm::device_uvector beta_sorted(0, stream); rmm::device_uvector X_active_cols(0, stream); auto execution_policy = handle.get_thrust_policy(); diff --git a/cpp/src/solver/solver.cu b/cpp/src/solver/solver.cu index 10d63d8650..b8ca76b20d 100644 --- a/cpp/src/solver/solver.cu +++ b/cpp/src/solver/solver.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -95,7 +95,7 @@ void sgdFit(raft::handle_t& handle, shuffle, tol, n_iter_no_change, - handle.get_stream()); + handle.get_stream().get()); } void sgdFit(raft::handle_t& handle, @@ -176,7 +176,7 @@ void sgdFit(raft::handle_t& handle, shuffle, tol, n_iter_no_change, - handle.get_stream()); + handle.get_stream().get()); } void sgdPredict(raft::handle_t& handle, @@ -200,7 +200,7 @@ void sgdPredict(raft::handle_t& handle, } sgdPredict( - handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream()); + handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream().get()); } void sgdPredict(raft::handle_t& handle, @@ -224,7 +224,7 @@ void sgdPredict(raft::handle_t& handle, } sgdPredict( - handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream()); + handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream().get()); } void sgdPredictBinaryClass(raft::handle_t& handle, @@ -248,7 +248,7 @@ void sgdPredictBinaryClass(raft::handle_t& handle, } sgdPredictBinaryClass( - handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream()); + handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream().get()); } void sgdPredictBinaryClass(raft::handle_t& handle, @@ -272,7 +272,7 @@ void sgdPredictBinaryClass(raft::handle_t& handle, } sgdPredictBinaryClass( - handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream()); + handle, input, n_rows, n_cols, coef, intercept, preds, loss_funct, handle.get_stream().get()); } int cdFit(raft::handle_t& handle, diff --git a/cpp/src/svm/kernelcache.cuh b/cpp/src/svm/kernelcache.cuh index 86f687368d..9883ca0f01 100644 --- a/cpp/src/svm/kernelcache.cuh +++ b/cpp/src/svm/kernelcache.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -383,7 +383,7 @@ class KernelCache { size_t kernel_tile_byte_limit = 1 << 30, size_t dense_extract_byte_limit = 1 << 30, bool is_precomputed = false) - : batch_cache(n_rows, cache_size, handle.get_stream()), + : batch_cache(n_rows, cache_size, handle.get_stream().get()), handle(handle), kernel(kernel), kernel_type(kernel_type), @@ -393,18 +393,18 @@ class KernelCache { n_ws(n_ws), svmType(svmType), is_precomputed(is_precomputed), - kernel_tile(0, handle.get_stream()), - matrix_l2(0, handle.get_stream()), - matrix_l2_ws(0, handle.get_stream()), - ws_idx_mod(n_ws, handle.get_stream()), - ws_idx_mod_svr(svmType == EPSILON_SVR ? n_ws : 0, handle.get_stream()), + kernel_tile(0, handle.get_stream().get()), + matrix_l2(0, handle.get_stream().get()), + matrix_l2_ws(0, handle.get_stream().get()), + ws_idx_mod(n_ws, handle.get_stream().get()), + ws_idx_mod_svr(svmType == EPSILON_SVR ? n_ws : 0, handle.get_stream().get()), x_ws_csr(nullptr), x_ws_dense(0, handle.get_stream()), indptr_batched(0, handle.get_stream()), ws_cache_idx(n_ws * 2, handle.get_stream()) { ASSERT(kernel != nullptr || is_precomputed, "Kernel pointer required for KernelCache!"); - stream = handle.get_stream(); + stream = handle.get_stream().get(); batching_enabled = false; is_csr = !isDenseType(); diff --git a/cpp/src/svm/linear.cu b/cpp/src/svm/linear.cu index 450553aaa8..8b64fd6b80 100644 --- a/cpp/src/svm/linear.cu +++ b/cpp/src/svm/linear.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include @@ -112,7 +112,7 @@ class WorkerHandle { : handle_ptr{new raft::handle_t{h.get_next_usable_stream(stream_id)}}, stream_id(stream_id), handle(*handle_ptr), - stream(h.get_next_usable_stream(stream_id)) + stream(h.get_next_usable_stream(stream_id).get()) { } @@ -140,7 +140,7 @@ int fit(const raft::handle_t& handle, ASSERT(nClasses > 1 && classes != nullptr, "Must have > 1 class for classification"); } - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); const int coefCols = nClasses <= 2 ? 1 : nClasses; const std::size_t coefRows = nCols + int(params.fit_intercept); diff --git a/cpp/src/svm/results.cuh b/cpp/src/svm/results.cuh index e3da603300..09eed26fd7 100644 --- a/cpp/src/svm/results.cuh +++ b/cpp/src/svm/results.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -67,7 +67,7 @@ class Results { SvmType svmType, bool is_precomputed = false) : rmm_alloc(rmm::mr::get_current_device_resource_ref()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), handle(handle), n_rows(n_rows), n_cols(n_cols), diff --git a/cpp/src/svm/smosolver.h b/cpp/src/svm/smosolver.h index fe13ffed9b..b13e9dc82c 100644 --- a/cpp/src/svm/smosolver.h +++ b/cpp/src/svm/smosolver.h @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -73,7 +73,7 @@ class SmoSolver { nochange_steps(param.nochange_steps), epsilon(param.epsilon), svmType(param.svmType), - stream(handle.get_stream()), + stream(handle.get_stream().get()), return_buff(2, stream), alpha(0, stream), C_vec(0, stream), diff --git a/cpp/src/svm/sparse_util.cuh b/cpp/src/svm/sparse_util.cuh index baa20832e4..0719434a75 100644 --- a/cpp/src/svm/sparse_util.cuh +++ b/cpp/src/svm/sparse_util.cuh @@ -438,21 +438,21 @@ void matrixRowNorm(const raft::handle_t& handle, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else if (norm == raft::linalg::NormType::L1Norm) { raft::linalg::rowNorm( target, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else if (norm == raft::linalg::NormType::LinfNorm) { raft::linalg::rowNorm( target, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else { RAFT_FAIL("Unsupported norm type"); } @@ -463,21 +463,21 @@ void matrixRowNorm(const raft::handle_t& handle, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else if (norm == raft::linalg::NormType::L1Norm) { raft::linalg::rowNorm( target, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else if (norm == raft::linalg::NormType::LinfNorm) { raft::linalg::rowNorm( target, matrix.data_handle(), matrix.extent(1), //! cols first arg! matrix.extent(0), - handle.get_stream()); + handle.get_stream().get()); } else { RAFT_FAIL("Unsupported norm type"); } @@ -628,7 +628,7 @@ void extractRows(raft::device_csr_matrix_view matrix_in, int num_indices, const raft::handle_t& handle) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto csr_struct_in = matrix_in.structure_view(); // initialize dense target @@ -733,7 +733,7 @@ void extractRows(raft::device_csr_matrix_view matrix_in, int num_indices, const raft::handle_t& handle) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto csr_struct_in = matrix_in.structure_view(); int* indptr_in = csr_struct_in.get_indptr().data(); int* indices_in = csr_struct_in.get_indices().data(); @@ -807,7 +807,7 @@ void extractRows(raft::device_csr_matrix_view matrix_in, int num_indices, const raft::handle_t& handle) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto csr_struct_in = matrix_in.structure_view(); int* indptr_in = csr_struct_in.get_indptr().data(); int* indices_in = csr_struct_in.get_indices().data(); diff --git a/cpp/src/svm/svc_impl.cuh b/cpp/src/svm/svc_impl.cuh index c73198c2d7..422d5751b1 100644 --- a/cpp/src/svm/svc_impl.cuh +++ b/cpp/src/svm/svc_impl.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -141,7 +141,7 @@ int svcFitX(const raft::handle_t& handle, // ML::detail::streamSyncer _(handle_impl.getImpl()); const raft::handle_t& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); { rmm::device_uvector unique_labels(0, stream); model.n_classes = raft::label::getUniquelabels(unique_labels, labels, n_rows, stream); @@ -252,7 +252,7 @@ void svcPredictX(const raft::handle_t& handle, } const raft::handle_t& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); rmm::device_uvector K(checked_mul(n_batch, model.n_support), stream); rmm::device_uvector y(n_rows, stream); @@ -480,7 +480,7 @@ void svcPredictSparse(const raft::handle_t& handle, template void svmFreeBuffers(const raft::handle_t& handle, SvmModel& m) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); rmm::device_async_resource_ref rmm_alloc = rmm::mr::get_current_device_resource_ref(); if (m.dual_coefs) rmm_alloc.deallocate(stream, m.dual_coefs, m.n_support * sizeof(math_t)); if (m.support_idx) rmm_alloc.deallocate(stream, m.support_idx, m.n_support * sizeof(int)); diff --git a/cpp/src/svm/svr_impl.cuh b/cpp/src/svm/svr_impl.cuh index 7113c5a4cf..9557177a9c 100644 --- a/cpp/src/svm/svr_impl.cuh +++ b/cpp/src/svm/svr_impl.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -53,7 +53,7 @@ int svrFitX(const raft::handle_t& handle, // ML::detail::streamSyncer _(handle_impl.getImpl()); const raft::handle_t& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); bool is_precomputed = kernel_params.kernel == ML::matrix::KernelType::PRECOMPUTED; diff --git a/cpp/src/tsa/auto_arima.cu b/cpp/src/tsa/auto_arima.cu index 72ddef6c98..bc660e3ca4 100644 --- a/cpp/src/tsa/auto_arima.cu +++ b/cpp/src/tsa/auto_arima.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include "auto_arima.cuh" @@ -15,7 +15,7 @@ int divide_by_mask_build_index(const raft::handle_t& handle, int* d_index, int batch_size) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); return ML::TimeSeries::divide_by_mask_build_index(d_mask, d_index, batch_size, stream); } @@ -29,7 +29,7 @@ inline void divide_by_mask_execute_helper(const raft::handle_t& handle, int batch_size, int n_obs) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::TimeSeries::divide_by_mask_execute( d_in, d_mask, d_index, d_out0, d_out1, batch_size, n_obs, stream); } @@ -79,7 +79,7 @@ inline void divide_by_min_build_index_helper(const raft::handle_t& handle, int batch_size, int n_sub) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::TimeSeries::divide_by_min_build_index( d_matrix, d_batch, d_index, h_size, batch_size, n_sub, stream); } @@ -116,7 +116,7 @@ inline void divide_by_min_execute_helper(const raft::handle_t& handle, int n_sub, int n_obs) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::TimeSeries::divide_by_min_execute( d_in, d_batch, d_index, hd_out, batch_size, n_sub, n_obs, stream); } @@ -165,7 +165,7 @@ void build_division_map(const raft::handle_t& handle, int batch_size, int n_sub) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::TimeSeries::build_division_map( hd_id, h_size, d_id_to_pos, d_id_to_model, batch_size, n_sub, stream); } @@ -180,7 +180,7 @@ inline void merge_series_helper(const raft::handle_t& handle, int n_sub, int n_obs) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::TimeSeries::merge_series( hd_in, d_id_to_pos, d_id_to_sub, d_out, batch_size, n_sub, n_obs, stream); } diff --git a/cpp/src/tsa/stationarity.cu b/cpp/src/tsa/stationarity.cu index b48e4a742f..de19e3835b 100644 --- a/cpp/src/tsa/stationarity.cu +++ b/cpp/src/tsa/stationarity.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -25,7 +25,7 @@ inline void kpss_test_helper(const raft::handle_t& handle, DataT pval_threshold) { const auto& handle_impl = handle; - cudaStream_t stream = handle_impl.get_stream(); + cudaStream_t stream = handle_impl.get_stream().get(); MLCommon::TimeSeries::kpss_test(d_y, results, batch_size, n_obs, d, D, s, stream, pval_threshold); } diff --git a/cpp/src/tsne/barnes_hut_tsne.cuh b/cpp/src/tsne/barnes_hut_tsne.cuh index 4da018c549..45cfbde3d5 100644 --- a/cpp/src/tsne/barnes_hut_tsne.cuh +++ b/cpp/src/tsne/barnes_hut_tsne.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -49,7 +49,7 @@ std::pair Barnes_Hut(value_t* VAL, const value_idx n, const TSNEParams& params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); value_t kl_div = 0; diff --git a/cpp/src/tsne/exact_kernels.cuh b/cpp/src/tsne/exact_kernels.cuh index 693a576598..ea86be71b7 100644 --- a/cpp/src/tsne/exact_kernels.cuh +++ b/cpp/src/tsne/exact_kernels.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -144,7 +144,7 @@ void perplexity_search(const value_t* restrict distances, const raft::handle_t& handle) { const float desired_entropy = logf(perplexity); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); if (dim == 2) sigmas_kernel_2d<<>>( diff --git a/cpp/src/tsne/exact_tsne.cuh b/cpp/src/tsne/exact_tsne.cuh index 4c5d28c1cf..eb09a7e86e 100644 --- a/cpp/src/tsne/exact_tsne.cuh +++ b/cpp/src/tsne/exact_tsne.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -41,7 +41,7 @@ std::pair Exact_TSNE(value_t* VAL, const value_idx n, const TSNEParams& params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); value_t kl_div = 0; const value_idx dim = params.dim; diff --git a/cpp/src/tsne/fft_tsne.cuh b/cpp/src/tsne/fft_tsne.cuh index 34adb310a8..e2b46d5a5f 100644 --- a/cpp/src/tsne/fft_tsne.cuh +++ b/cpp/src/tsne/fft_tsne.cuh @@ -174,7 +174,7 @@ std::pair FFT_TSNE(value_t* VAL, const value_idx n, const TSNEParams& params) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); auto thrust_policy = handle.get_thrust_policy(); // Fixed seeds use deterministic accumulation paths; unseeded runs keep the // original faster atomic/reduction paths. diff --git a/cpp/src/tsne/tsne_runner.cuh b/cpp/src/tsne/tsne_runner.cuh index b7f498ee01..5f432108ef 100644 --- a/cpp/src/tsne/tsne_runner.cuh +++ b/cpp/src/tsne/tsne_runner.cuh @@ -56,7 +56,7 @@ class TSNE_runner { input(input_), k_graph(k_graph_), params(params_), - COO_Matrix(handle_.get_stream()) + COO_Matrix(handle_.get_stream().get()) { this->n = input.n; this->p = input.d; @@ -89,7 +89,7 @@ class TSNE_runner { "# of Nearest Neighbors should be at least 3 * perplexity. Your results" " might be a bit strange..."); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); const value_idx dim = params.dim; if (params.init == TSNE_INIT::RANDOM) { @@ -189,7 +189,7 @@ class TSNE_runner { // Get distances CUML_LOG_DEBUG("Getting distances."); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_uvector indices(0, stream); rmm::device_uvector distances(0, stream); diff --git a/cpp/src/umap/init_embed/runner.cuh b/cpp/src/umap/init_embed/runner.cuh index 7e611d4d89..da93c93f2a 100644 --- a/cpp/src/umap/init_embed/runner.cuh +++ b/cpp/src/umap/init_embed/runner.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -32,12 +32,14 @@ void run(const raft::handle_t& handle, /** * Initial algo uses FAISS indices */ - case 0: RandomInit::launcher(n, d, params, embedding, handle.get_stream()); break; + case 0: + RandomInit::launcher(n, d, params, embedding, handle.get_stream().get()); + break; case 1: try { SpectralInit::launcher(handle, n, d, coo, params, embedding); } catch (const raft::exception& e) { CUML_LOG_WARN("Spectral initialization failed, using random initialization instead."); - RandomInit::launcher(n, d, params, embedding, handle.get_stream()); + RandomInit::launcher(n, d, params, embedding, handle.get_stream().get()); } break; diff --git a/cpp/src/umap/init_embed/spectral_algo.cuh b/cpp/src/umap/init_embed/spectral_algo.cuh index 969a5397bc..c2ead128d2 100644 --- a/cpp/src/umap/init_embed/spectral_algo.cuh +++ b/cpp/src/umap/init_embed/spectral_algo.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -40,7 +40,7 @@ void launcher(const raft::handle_t& handle, UMAPParams* params, T* embedding) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ASSERT(n > static_cast(params->n_components), "Spectral layout requires n_samples > n_components"); diff --git a/cpp/src/umap/runner.cuh b/cpp/src/umap/runner.cuh index d3bc5d6721..943b7038a4 100644 --- a/cpp/src/umap/runner.cuh +++ b/cpp/src/umap/runner.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -162,7 +162,7 @@ void _get_strengths(const raft::handle_t& handle, value_t* out_sigmas = nullptr, value_t* out_rhos = nullptr) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int n_neighbors = params->n_neighbors; nnz_t n_x_n_neighbors = static_cast(inputs.n) * n_neighbors; @@ -211,7 +211,7 @@ void _get_graph(const raft::handle_t& handle, value_t* out_rhos = nullptr) { raft::common::nvtx::range fun_scope("umap::supervised::_get_graph"); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::default_logger().set_level(params->verbosity); @@ -238,7 +238,7 @@ void _get_graph_supervised(const raft::handle_t& handle, { if (params->target_n_neighbors == -1) params->target_n_neighbors = params->n_neighbors; - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); /* Nested scopes used here to drop resources earlier, reducing device memory usage */ raft::sparse::COO ci_graph(stream); @@ -270,7 +270,7 @@ void _refine(const raft::handle_t& handle, raft::sparse::COO* graph, value_t* embeddings) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::default_logger().set_level(params->verbosity); int n_epochs = get_n_epochs(params, inputs.n); @@ -290,7 +290,7 @@ void _init_and_refine(const raft::handle_t& handle, raft::sparse::COO* graph, value_t* embeddings) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::default_logger().set_level(params->verbosity); int n_epochs = get_n_epochs(params, inputs.n); @@ -321,7 +321,7 @@ void _fit(const raft::handle_t& handle, int n_epochs = get_n_epochs(params, inputs.n); - raft::sparse::COO graph(stream); + raft::sparse::COO graph(stream.get()); UMAPAlgo::_get_graph( handle, inputs, params, &graph, out_sigmas, out_rhos); @@ -340,7 +340,7 @@ void _fit(const raft::handle_t& handle, */ raft::common::nvtx::push_range("umap::embedding"); InitEmbed::run( - handle, inputs.n, inputs.d, &graph, params, embeddings_ptr, stream, params->init); + handle, inputs.n, inputs.d, &graph, params, embeddings_ptr, stream.get(), params->init); if (params->callback) { params->callback->setup(inputs.n, params->n_components); @@ -351,7 +351,7 @@ void _fit(const raft::handle_t& handle, * Run simplicial set embedding to approximate low-dimensional representation */ SimplSetEmbed::run( - inputs.n, inputs.d, &graph, params, embeddings_ptr, n_epochs, stream); + inputs.n, inputs.d, &graph, params, embeddings_ptr, n_epochs, stream.get()); raft::common::nvtx::pop_range(); if (params->callback) params->callback->on_train_end(embeddings_ptr); @@ -368,7 +368,7 @@ void _fit_supervised(const raft::handle_t& handle, { raft::common::nvtx::range fun_scope("umap::supervised::fit"); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); ML::default_logger().set_level(params->verbosity); int n_epochs = get_n_epochs(params, inputs.n); @@ -424,7 +424,7 @@ void _transform(const raft::handle_t& handle, value_t* transformed) { raft::common::nvtx::range fun_scope("umap::transform"); - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); ML::default_logger().set_level(params->verbosity); diff --git a/cpp/src/umap/umap.cu b/cpp/src/umap/umap.cu index 0dc5f939fa..5ff6466262 100644 --- a/cpp/src/umap/umap.cu +++ b/cpp/src/umap/umap.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,7 +12,7 @@ namespace UMAP { void find_ab(const raft::handle_t& handle, UMAPParams* params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); UMAPAlgo::find_ab(params, stream); } diff --git a/cpp/src/umap/umap.cuh b/cpp/src/umap/umap.cuh index 1483c76ab3..3c49703024 100644 --- a/cpp/src/umap/umap.cuh +++ b/cpp/src/umap/umap.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -43,7 +43,7 @@ inline std::unique_ptr> _get_graph( float* knn_dists, // precomputed distances UMAPParams* params) { - auto graph = std::make_unique>(handle.get_stream()); + auto graph = std::make_unique>(handle.get_stream().get()); if (knn_indices != nullptr && knn_dists != nullptr) { CUML_LOG_DEBUG("Calling UMAP::get_graph() with precomputed KNN"); @@ -274,7 +274,7 @@ inline void _inverse_transform(const raft::handle_t& handle, UMAPParams* params, int n_epochs) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); // Compute epochs_per_sample from graph weights rmm::device_uvector epochs_per_sample(nnz, stream); diff --git a/cpp/src_prims/selection/knn.cuh b/cpp/src_prims/selection/knn.cuh index b6687b8c92..54f10b09af 100644 --- a/cpp/src_prims/selection/knn.cuh +++ b/cpp/src_prims/selection/knn.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -192,7 +192,7 @@ void class_probs(const raft::handle_t& handle, const float* weights = nullptr) { for (std::size_t i = 0; i < y.size(); i++) { - cudaStream_t stream = handle.get_next_usable_stream(); + cudaStream_t stream = handle.get_next_usable_stream().get(); int n_unique_labels = n_unique[i]; size_t cur_size = n_query_rows * n_unique_labels; @@ -274,7 +274,7 @@ void knn_classify(const raft::handle_t& handle, for (std::size_t i = 0; i < n_unique.size(); i++) { int size = n_unique[i]; - cudaStream_t stream = handle.get_next_usable_stream(i); + cudaStream_t stream = handle.get_next_usable_stream(i).get(); tmp_probs.emplace_back(n_query_rows * size, stream); probs.push_back(tmp_probs.back().data()); @@ -293,7 +293,7 @@ void knn_classify(const raft::handle_t& handle, dim3 blk(TPB_X, 1, 1); for (std::size_t i = 0; i < y.size(); i++) { - cudaStream_t stream = handle.get_next_usable_stream(i); + cudaStream_t stream = handle.get_next_usable_stream(i).get(); int n_unique_labels = n_unique[i]; @@ -347,7 +347,7 @@ void knn_regress(const raft::handle_t& handle, * Vote average regression value */ for (std::size_t i = 0; i < y.size(); i++) { - cudaStream_t stream = handle.get_next_usable_stream(); + cudaStream_t stream = handle.get_next_usable_stream().get(); regress_avg_kernel <<(TPB_X)), TPB_X, 0, stream>>>( diff --git a/cpp/tests/mg/rf_quantile_test.cu b/cpp/tests/mg/rf_quantile_test.cu index 3a39651a24..364a97c5e6 100644 --- a/cpp/tests/mg/rf_quantile_test.cu +++ b/cpp/tests/mg/rf_quantile_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,6 +12,8 @@ #include #include +#include + #include #include #include @@ -117,7 +119,7 @@ class RfMgQuantileTest : public ::testing::Test { RAFT_CUDA_TRY(cudaSetDevice(local_rank)); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); raft::comms::initialize_mpi_comms(&handle, MPI_COMM_WORLD); constexpr int n_cols = 3; @@ -157,7 +159,7 @@ class RfMgQuantileTest : public ::testing::Test { std::vector h_reference_quantiles(static_cast(n_cols) * max_n_bins); if (rank == 0) { - raft::handle_t reference_handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t reference_handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); rmm::device_uvector reference_data(h_global_data.size(), reference_handle.get_stream()); raft::update_device(reference_data.data(), h_global_data.data(), diff --git a/cpp/tests/mg/rf_test.cu b/cpp/tests/mg/rf_test.cu index 69a464a599..8adc1d8f44 100644 --- a/cpp/tests/mg/rf_test.cu +++ b/cpp/tests/mg/rf_test.cu @@ -15,6 +15,8 @@ #include #include +#include + #include #include @@ -289,7 +291,7 @@ class RfMgPropertyTestImpl { RAFT_CUDA_TRY(cudaSetDevice(local_rank)); auto stream_pool = std::make_shared(params.handle_n_streams); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); raft::comms::initialize_mpi_comms(&handle, MPI_COMM_WORLD); auto local_rows = local_rows_for_rank(params.n_rows, rank, size, params.partition_kind); @@ -369,7 +371,8 @@ class RfMgPropertyTestImpl { params, global_rows, h_global_X, h_global_y, h_global_sample_weights); auto single_node_stream_pool = std::make_shared(params.handle_n_streams); - raft::handle_t single_node_handle(rmm::cuda_stream_per_thread, single_node_stream_pool); + raft::handle_t single_node_handle(cuda::stream_ref{cudaStreamPerThread}, + single_node_stream_pool); rmm::device_uvector global_X(h_global_X.size(), single_node_handle.get_stream()); rmm::device_uvector global_y(h_global_y.size(), single_node_handle.get_stream()); rmm::device_uvector global_sample_weights(h_global_sample_weights.size(), diff --git a/cpp/tests/prims/fillna.cu b/cpp/tests/prims/fillna.cu index 122246f088..74c2d69643 100644 --- a/cpp/tests/prims/fillna.cu +++ b/cpp/tests/prims/fillna.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -78,7 +78,7 @@ class FillnaTest : public ::testing::TestWithParam> { handle.sync_stream(handle.get_stream()); /* Compute using tested prims */ - fillna(y.data(), params.batch_size, params.n_obs, handle.get_stream()); + fillna(y.data(), params.batch_size, params.n_obs, handle.get_stream().get()); /* Compute reference results. * Note: this is done with a sliding window: we find ranges of missing @@ -125,7 +125,7 @@ class FillnaTest : public ::testing::TestWithParam> { y.data(), params.n_obs * params.batch_size, MLCommon::CompareApprox(params.tolerance), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } diff --git a/cpp/tests/prims/hinge.cu b/cpp/tests/prims/hinge.cu index e3b7b27900..b5e6d5ba79 100644 --- a/cpp/tests/prims/hinge.cu +++ b/cpp/tests/prims/hinge.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -27,7 +27,7 @@ class HingeLossTest : public ::testing::TestWithParam> { public: HingeLossTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), in(params.len, stream), out(1, stream), out_lasso(1, stream), diff --git a/cpp/tests/prims/jones_transform.cu b/cpp/tests/prims/jones_transform.cu index 161158519e..631d92822e 100644 --- a/cpp/tests/prims/jones_transform.cu +++ b/cpp/tests/prims/jones_transform.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. * + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include "test_utils.h" @@ -34,7 +34,7 @@ template public: JonesTransTest() : params(::testing::TestWithParam::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), nElements(params.batchSize * params.pValue), d_golden_ar_trans(0, stream), d_computed_ar_trans(0, stream), diff --git a/cpp/tests/prims/knn_classify.cu b/cpp/tests/prims/knn_classify.cu index fd6ed645c4..1743a3b5c6 100644 --- a/cpp/tests/prims/knn_classify.cu +++ b/cpp/tests/prims/knn_classify.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -34,7 +34,7 @@ class KNNClassifyTest : public ::testing::TestWithParam { public: KNNClassifyTest() : params(::testing::TestWithParam::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), train_samples(params.rows * params.cols, stream), train_labels(params.rows, stream), pred_labels(params.rows, stream), diff --git a/cpp/tests/prims/knn_regression.cu b/cpp/tests/prims/knn_regression.cu index 4ff7abcc46..2bf5266a7a 100644 --- a/cpp/tests/prims/knn_regression.cu +++ b/cpp/tests/prims/knn_regression.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -72,7 +72,7 @@ class KNNRegressionTest : public ::testing::TestWithParam { public: KNNRegressionTest() : params(::testing::TestWithParam::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), train_samples(params.rows * params.cols, stream), train_labels(params.rows, stream), pred_labels(params.rows, stream), diff --git a/cpp/tests/prims/kselection.cu b/cpp/tests/prims/kselection.cu index 579e8193da..cc4f2771ed 100644 --- a/cpp/tests/prims/kselection.cu +++ b/cpp/tests/prims/kselection.cu @@ -1,11 +1,13 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include #include +#include + #include #include #include @@ -72,7 +74,8 @@ template for (int rIndex = 0; rIndex < rows; rIndex++) { // input data TypeV* h_arr = new TypeV[N]; - raft::update_host(h_arr, d_arr + rIndex * N, N, rmm::cuda_stream_default); + raft::update_host( + h_arr, d_arr + rIndex * N, N, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); KVPair* topk = new KVPair[N]; for (int j = 0; j < N; j++) { topk[j].val = h_arr[j]; @@ -80,9 +83,11 @@ template } // result reference TypeV* h_outv = new TypeV[k]; - raft::update_host(h_outv, d_outv + rIndex * k, k, rmm::cuda_stream_default); + raft::update_host( + h_outv, d_outv + rIndex * k, k, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); TypeK* h_outk = new TypeK[k]; - raft::update_host(h_outk, d_outk + rIndex * k, k, rmm::cuda_stream_default); + raft::update_host( + h_outk, d_outk + rIndex * k, k, cuda::stream_ref{cudaStream_t{cudaStreamDefault}}); // calculate the result partSortKVPair(topk, N, k); diff --git a/cpp/tests/prims/linalg_block.cu b/cpp/tests/prims/linalg_block.cu index b180b4f9f8..2e2ab34eca 100644 --- a/cpp/tests/prims/linalg_block.cu +++ b/cpp/tests/prims/linalg_block.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -71,7 +71,7 @@ class BlockGemmTest : public ::testing::TestWithParam> { rmm::device_uvector a(params.m * params.k * params.batch_size, handle.get_stream()); rmm::device_uvector b(params.k * params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector c(params.m * params.n * params.batch_size, handle.get_stream()); + rmm::device_uvector c(params.m * params.n * params.batch_size, handle.get_stream().get()); std::vector h_a(params.m * params.k * params.batch_size); std::vector h_b(params.k * params.n * params.batch_size); @@ -79,8 +79,10 @@ class BlockGemmTest : public ::testing::TestWithParam> { /* Generate random data on device */ raft::random::Rng r(params.seed); - r.uniform(a.data(), params.m * params.k * params.batch_size, (T)-2, (T)2, handle.get_stream()); - r.uniform(b.data(), params.k * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); + r.uniform( + a.data(), params.m * params.k * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); + r.uniform( + b.data(), params.k * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); /* Generate random alpha */ std::default_random_engine generator(params.seed); @@ -89,22 +91,22 @@ class BlockGemmTest : public ::testing::TestWithParam> { /* Copy to host */ raft::update_host( - h_a.data(), a.data(), params.m * params.k * params.batch_size, handle.get_stream()); + h_a.data(), a.data(), params.m * params.k * params.batch_size, handle.get_stream().get()); raft::update_host( - h_b.data(), b.data(), params.k * params.n * params.batch_size, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + h_b.data(), b.data(), params.k * params.n * params.batch_size, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); /* Compute using tested prims */ block_gemm_test_kernel - <<>>(params.transa, - params.transb, - params.m, - params.n, - params.k, - alpha, - a.data(), - b.data(), - c.data()); + <<>>(params.transa, + params.transb, + params.m, + params.n, + params.k, + alpha, + a.data(), + b.data(), + c.data()); /* Compute reference results */ for (int bid = 0; bid < params.batch_size; bid++) { @@ -129,7 +131,7 @@ class BlockGemmTest : public ::testing::TestWithParam> { c.data(), params.m * params.n * params.batch_size, MLCommon::CompareApprox(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -297,7 +299,7 @@ class BlockGemvTest : public ::testing::TestWithParam> { rmm::device_uvector a(params.m * params.n * params.batch_size, handle.get_stream()); rmm::device_uvector x(params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector y(params.m * params.batch_size, handle.get_stream()); + rmm::device_uvector y(params.m * params.batch_size, handle.get_stream().get()); std::vector h_a(params.m * params.n * params.batch_size); std::vector h_x(params.n * params.batch_size); @@ -305,8 +307,9 @@ class BlockGemvTest : public ::testing::TestWithParam> { /* Generate random data on device */ raft::random::Rng r(params.seed); - r.uniform(a.data(), params.m * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); - r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); + r.uniform( + a.data(), params.m * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); + r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); /* Generate random alpha */ std::default_random_engine generator(params.seed); @@ -315,14 +318,15 @@ class BlockGemvTest : public ::testing::TestWithParam> { /* Copy to host */ raft::update_host( - h_a.data(), a.data(), params.m * params.n * params.batch_size, handle.get_stream()); - raft::update_host(h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + h_a.data(), a.data(), params.m * params.n * params.batch_size, handle.get_stream().get()); + raft::update_host( + h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); /* Compute using tested prims */ int shared_mem_size = params.n * sizeof(T); block_gemv_test_kernel - <<>>( + <<>>( params.m, params.n, alpha, a.data(), x.data(), y.data(), params.preload); /* Compute reference results */ @@ -341,7 +345,7 @@ class BlockGemvTest : public ::testing::TestWithParam> { y.data(), params.m * params.batch_size, MLCommon::CompareApprox(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -434,7 +438,7 @@ class BlockDotTest : public ::testing::TestWithParam> { rmm::device_uvector x(params.n * params.batch_size, handle.get_stream()); rmm::device_uvector y(params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector dot_dev(params.batch_size, handle.get_stream()); + rmm::device_uvector dot_dev(params.batch_size, handle.get_stream().get()); std::vector h_x(params.n * params.batch_size); std::vector h_y(params.n * params.batch_size); @@ -442,23 +446,25 @@ class BlockDotTest : public ::testing::TestWithParam> { /* Generate random data on device */ raft::random::Rng r(params.seed); - r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); - r.uniform(y.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); + r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); + r.uniform(y.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); /* Copy to host */ - raft::update_host(h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream()); - raft::update_host(h_y.data(), y.data(), params.n * params.batch_size, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + raft::update_host( + h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream().get()); + raft::update_host( + h_y.data(), y.data(), params.n * params.batch_size, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); /* Compute using tested prims */ constexpr int BlockSize = 64; if (params.broadcast) block_dot_test_kernel - <<>>( + <<>>( params.n, x.data(), y.data(), dot_dev.data()); else block_dot_test_kernel - <<>>( + <<>>( params.n, x.data(), y.data(), dot_dev.data()); /* Compute reference results */ @@ -473,7 +479,7 @@ class BlockDotTest : public ::testing::TestWithParam> { dot_dev.data(), params.batch_size, MLCommon::CompareApprox(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -562,7 +568,7 @@ class BlockXaxtTest : public ::testing::TestWithParam> { rmm::device_uvector x(params.n * params.batch_size, handle.get_stream()); rmm::device_uvector A(params.n * params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector res_dev(params.batch_size, handle.get_stream()); + rmm::device_uvector res_dev(params.batch_size, handle.get_stream().get()); std::vector h_x(params.n * params.batch_size); std::vector h_A(params.n * params.n * params.batch_size); @@ -570,25 +576,27 @@ class BlockXaxtTest : public ::testing::TestWithParam> { /* Generate random data on device */ raft::random::Rng r(params.seed); - r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); - r.uniform(A.data(), params.n * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); + r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); + r.uniform( + A.data(), params.n * params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); /* Copy to host */ - raft::update_host(h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream()); raft::update_host( - h_A.data(), A.data(), params.n * params.n * params.batch_size, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream().get()); + raft::update_host( + h_A.data(), A.data(), params.n * params.n * params.batch_size, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); /* Compute using tested prims */ constexpr int BlockSize = 64; int shared_mem_size = params.n * sizeof(T); if (params.broadcast) block_xAxt_test_kernel - <<>>( + <<>>( params.n, x.data(), A.data(), res_dev.data(), params.preload); else block_xAxt_test_kernel - <<>>( + <<>>( params.n, x.data(), A.data(), res_dev.data(), params.preload); /* Compute reference results */ @@ -607,7 +615,7 @@ class BlockXaxtTest : public ::testing::TestWithParam> { res_dev.data(), params.batch_size, MLCommon::CompareApprox(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -673,15 +681,15 @@ class BlockAxTest : public ::testing::TestWithParam> { params = ::testing::TestWithParam>::GetParam(); - rmm::device_uvector x(params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector y(params.n * params.batch_size, handle.get_stream()); + rmm::device_uvector x(params.n * params.batch_size, handle.get_stream().get()); + rmm::device_uvector y(params.n * params.batch_size, handle.get_stream().get()); std::vector h_x(params.n * params.batch_size); std::vector h_y_ref(params.n * params.batch_size, (T)0); /* Generate random data on device */ raft::random::Rng r(params.seed); - r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream()); + r.uniform(x.data(), params.n * params.batch_size, (T)-2, (T)2, handle.get_stream().get()); /* Generate random alpha */ std::default_random_engine generator(params.seed); @@ -689,12 +697,13 @@ class BlockAxTest : public ::testing::TestWithParam> { T alpha = distribution(generator); /* Copy to host */ - raft::update_host(h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + raft::update_host( + h_x.data(), x.data(), params.n * params.batch_size, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); /* Compute using tested prims */ constexpr int BlockSize = 64; - block_ax_test_kernel<<>>( + block_ax_test_kernel<<>>( params.n, alpha, x.data(), y.data()); /* Compute reference results */ @@ -709,7 +718,7 @@ class BlockAxTest : public ::testing::TestWithParam> { y.data(), params.n * params.batch_size, MLCommon::CompareApprox(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -771,8 +780,9 @@ class BlockCovStabilityTest : public ::testing::TestWithParam>::GetParam(); - rmm::device_uvector d_in(params.n * params.n * params.batch_size, handle.get_stream()); - rmm::device_uvector d_out(params.n * params.n * params.batch_size, handle.get_stream()); + rmm::device_uvector d_in(params.n * params.n * params.batch_size, handle.get_stream().get()); + rmm::device_uvector d_out(params.n * params.n * params.batch_size, + handle.get_stream().get()); std::vector h_in(params.n * params.n * params.batch_size); std::vector h_out(params.n * params.n * params.batch_size); @@ -780,16 +790,16 @@ class BlockCovStabilityTest : public ::testing::TestWithParam - <<>>( + <<>>( params.n, d_in.data(), d_out.data()); /* Compute reference results */ @@ -813,7 +823,7 @@ class BlockCovStabilityTest : public ::testing::TestWithParam(params.eps), - handle.get_stream()); + handle.get_stream().get()); } void SetUp() override { basicTest(); } diff --git a/cpp/tests/prims/linearReg.cu b/cpp/tests/prims/linearReg.cu index 3dd2b93f0b..e4e318dc65 100644 --- a/cpp/tests/prims/linearReg.cu +++ b/cpp/tests/prims/linearReg.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -27,7 +27,7 @@ class LinRegLossTest : public ::testing::TestWithParam> { public: LinRegLossTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), in(params.len, stream), out(1, stream), out_lasso(1, stream), diff --git a/cpp/tests/prims/logisticReg.cu b/cpp/tests/prims/logisticReg.cu index 6fed36cd6d..3ac121edab 100644 --- a/cpp/tests/prims/logisticReg.cu +++ b/cpp/tests/prims/logisticReg.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -27,7 +27,7 @@ class LogRegLossTest : public ::testing::TestWithParam> { public: LogRegLossTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), in(params.len, stream), out(1, stream), out_lasso(1, stream), diff --git a/cpp/tests/prims/penalty.cu b/cpp/tests/prims/penalty.cu index 4243bdb6ca..f12c1d7c38 100644 --- a/cpp/tests/prims/penalty.cu +++ b/cpp/tests/prims/penalty.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -25,7 +25,7 @@ class PenaltyTest : public ::testing::TestWithParam> { public: PenaltyTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), in(params.len, stream), out_lasso(1, stream), out_ridge(1, stream), diff --git a/cpp/tests/sg/cd_test.cu b/cpp/tests/sg/cd_test.cu index 5b5ab91c58..e4177d4336 100644 --- a/cpp/tests/sg/cd_test.cu +++ b/cpp/tests/sg/cd_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -32,7 +32,7 @@ class CdTest : public ::testing::TestWithParam> { public: CdTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), data(params.n_row * params.n_col, stream), labels(params.n_row, stream), sample_weight(params.n_row, stream), diff --git a/cpp/tests/sg/dbscan_test.cu b/cpp/tests/sg/dbscan_test.cu index adec73a819..5c3e8b5254 100644 --- a/cpp/tests/sg/dbscan_test.cu +++ b/cpp/tests/sg/dbscan_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -115,9 +115,10 @@ class DbscanTest : public ::testing::TestWithParam> { score = adjusted_rand_index(handle, labels_ref.data(), labels.data(), params.n_row); if (score < 1.0) { - auto str = raft::arr2Str(labels_ref.data(), params.n_row, "labels_ref", handle.get_stream()); + auto str = + raft::arr2Str(labels_ref.data(), params.n_row, "labels_ref", handle.get_stream().get()); CUML_LOG_DEBUG("y: %s", str.c_str()); - str = raft::arr2Str(labels.data(), params.n_row, "labels", handle.get_stream()); + str = raft::arr2Str(labels.data(), params.n_row, "labels", handle.get_stream().get()); CUML_LOG_DEBUG("y_hat: %s", str.c_str()); CUML_LOG_DEBUG("Score = %lf", score); } @@ -270,7 +271,7 @@ class Dbscan2DSimple : public ::testing::TestWithParam> { if (eps_nn_method == Dbscan::EpsNnMethod::RBC) { std::cout << "RBC test" << std::endl; } raft::handle_t handle; - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); params = ::testing::TestWithParam>::GetParam(); diff --git a/cpp/tests/sg/hdbscan_test.cu b/cpp/tests/sg/hdbscan_test.cu index e1df28037a..61eea9b572 100644 --- a/cpp/tests/sg/hdbscan_test.cu +++ b/cpp/tests/sg/hdbscan_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2021-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -101,10 +101,10 @@ class HDBSCANTest : public ::testing::TestWithParam> { out, core_dists.data()); - handle.sync_stream(handle.get_stream()); + handle.sync_stream(handle.get_stream().get()); score = raft::stats::adjusted_rand_index( - out.get_labels(), labels_ref.data(), params.n_row, handle.get_stream()); + out.get_labels(), labels_ref.data(), params.n_row, handle.get_stream().get()); if (score < 0.85) { std::cout << "Test failed. score=" << score << std::endl; @@ -306,19 +306,20 @@ class ClusterSelectionTest : public ::testing::TestWithParam(0), params.cluster_selection_epsilon); - handle.sync_stream(handle.get_stream()); + handle.sync_stream(handle.get_stream().get()); ASSERT_TRUE(MLCommon::devArrMatch(probabilities.data(), params.probabilities.data(), params.n_row, MLCommon::CompareApprox(1e-4), - handle.get_stream())); + handle.get_stream().get())); - rmm::device_uvector labels_ref(params.n_row, handle.get_stream()); - raft::update_device(labels_ref.data(), params.labels.data(), params.n_row, handle.get_stream()); + rmm::device_uvector labels_ref(params.n_row, handle.get_stream().get()); + raft::update_device( + labels_ref.data(), params.labels.data(), params.n_row, handle.get_stream().get()); score = raft::stats::adjusted_rand_index( - labels.data(), labels_ref.data(), params.n_row, handle.get_stream()); - handle.sync_stream(handle.get_stream()); + labels.data(), labels_ref.data(), params.n_row, handle.get_stream().get()); + handle.sync_stream(handle.get_stream().get()); } void SetUp() override { basicTest(); } @@ -459,7 +460,7 @@ class AllPointsMembershipVectorsTest params.expected_probabilities.data(), params.n_row * n_selected_clusters, MLCommon::CompareApprox(1e-5), - handle.get_stream())); + handle.get_stream().get())); } void SetUp() override { basicTest(); } @@ -554,11 +555,11 @@ class ApproximatePredictTest : public ::testing::TestWithParam(0), params.cluster_selection_epsilon); - rmm::device_uvector core_dists{static_cast(params.n_row), handle.get_stream()}; + rmm::device_uvector core_dists{static_cast(params.n_row), handle.get_stream().get()}; ML::HDBSCAN::Common::PredictionData pred_data( handle, params.n_row, params.n_col, core_dists.data()); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_uvector mutual_reachability_indptr(params.n_row + 1, stream); raft::sparse::COO mutual_reachability_coo(stream, (params.min_samples + 1) * params.n_row * 2); @@ -645,20 +646,20 @@ class ApproximatePredictTest : public ::testing::TestWithParam(), - handle.get_stream())); + handle.get_stream().get())); ASSERT_TRUE(MLCommon::devArrMatch(out_probabilities.data(), params.expected_probabilities.data(), params.n_points_to_predict, MLCommon::CompareApprox(1e-2), - handle.get_stream())); + handle.get_stream().get())); } void SetUp() override { basicTest(); } @@ -754,13 +755,13 @@ class MembershipVectorTest : public ::testing::TestWithParam membership_vec(params.n_points_to_predict * n_selected_clusters, - handle.get_stream()); + handle.get_stream().get()); - rmm::device_uvector core_dists{static_cast(params.n_row), handle.get_stream()}; + rmm::device_uvector core_dists{static_cast(params.n_row), handle.get_stream().get()}; ML::HDBSCAN::Common::PredictionData prediction_data_( handle, params.n_row, params.n_col, core_dists.data()); - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_uvector mutual_reachability_indptr(params.n_row + 1, stream); raft::sparse::COO mutual_reachability_coo(stream, (params.min_samples + 1) * params.n_row * 2); @@ -846,7 +847,7 @@ class MembershipVectorTest : public ::testing::TestWithParam(1e-4), - handle.get_stream())); + handle.get_stream().get())); } void SetUp() override { basicTest(); } diff --git a/cpp/tests/sg/holtwinters_test.cu b/cpp/tests/sg/holtwinters_test.cu index 3053eef34e..8fcc32de42 100644 --- a/cpp/tests/sg/holtwinters_test.cu +++ b/cpp/tests/sg/holtwinters_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -41,7 +41,7 @@ class HoltWintersTest : public ::testing::TestWithParam> { public: HoltWintersTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), level_ptr(0, stream), trend_ptr(0, stream), season_ptr(0, stream), diff --git a/cpp/tests/sg/isolation_forest_test.cu b/cpp/tests/sg/isolation_forest_test.cu index 2c2b1479c2..2355813bf8 100644 --- a/cpp/tests/sg/isolation_forest_test.cu +++ b/cpp/tests/sg/isolation_forest_test.cu @@ -28,6 +28,7 @@ #include #include +#include #include #include #include @@ -103,8 +104,8 @@ class IsolationForestTest : public ::testing::Test { void SetUp() override { stream_pool = std::make_shared(4); - handle = std::make_unique(rmm::cuda_stream_per_thread, stream_pool); - stream = handle->get_stream(); + handle = std::make_unique(cuda::stream_ref{cudaStreamPerThread}, stream_pool); + stream = handle->get_stream().get(); } void TearDown() override diff --git a/cpp/tests/sg/knn_test.cu b/cpp/tests/sg/knn_test.cu index 4d6c0f2c00..7cd75ab636 100644 --- a/cpp/tests/sg/knn_test.cu +++ b/cpp/tests/sg/knn_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -71,7 +71,7 @@ void create_index_parts(raft::handle_t& handle, const KNNInputs& params, const float* centers) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); gen_blobs(handle, query_data, query_labels, @@ -120,7 +120,7 @@ class KNNTest : public ::testing::TestWithParam { public: KNNTest() : params(::testing::TestWithParam::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), index_data(params.n_rows * params.n_cols * params.n_parts, stream), index_labels(params.n_rows * params.n_parts, stream), search_data(params.n_query_row * params.n_cols, stream), @@ -275,7 +275,7 @@ class KNNTest : public ::testing::TestWithParam { private: void create_data() { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); rmm::device_uvector rand_centers(params.n_centers * params.n_cols, stream); Rng r(0, GeneratorType::GenPhilox); diff --git a/cpp/tests/sg/lars_test.cu b/cpp/tests/sg/lars_test.cu index b0b24bd0fb..8f36d2a2a2 100644 --- a/cpp/tests/sg/lars_test.cu +++ b/cpp/tests/sg/lars_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -45,7 +45,7 @@ class LarsTest : public ::testing::Test { void testSelectMostCorrelated() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); math_t cj; int idx; rmm::device_uvector workspace(n_cols, stream); @@ -57,7 +57,7 @@ class LarsTest : public ::testing::Test { void testMoveToActive() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); ML::Solver::Lars::moveToActive(handle.get_cublas_handle(), &n_active, 3, @@ -101,7 +101,7 @@ class LarsTest : public ::testing::Test { void calcUExp(math_t* G, int n_cols, math_t* U_dev_exp) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_scalar devInfo(stream); rmm::device_uvector workspace(0, stream); int n_work; @@ -127,7 +127,7 @@ class LarsTest : public ::testing::Test { // Initialize a mix of G and U matrices to test updateCholesky void initGU(math_t* GU, math_t* G, math_t* U, int n_active, bool copy_G) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); const int ld_U = n_cols; // First we copy over all elements, because the factorization only replaces // the upper triangular part. This way it will be easier to compare to the @@ -145,7 +145,7 @@ class LarsTest : public ::testing::Test { void testUpdateCholesky() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); const int ld_X = n_rows; const int ld_G = n_cols; const int ld_U = ld_G; @@ -214,7 +214,7 @@ class LarsTest : public ::testing::Test { void testCalcW0() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); n_active = 4; const int ld_U = n_cols; rmm::device_uvector ws(n_active, stream); @@ -229,7 +229,7 @@ class LarsTest : public ::testing::Test { void testCalcA() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); n_active = 4; rmm::device_uvector ws(n_active, stream); raft::update_device(ws.data(), ws0_exp, n_active, stream); @@ -241,7 +241,7 @@ class LarsTest : public ::testing::Test { void testEquiangular() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); n_active = 4; rmm::device_uvector workspace(0, stream); rmm::device_uvector u_eq(n_rows, stream); @@ -301,7 +301,7 @@ class LarsTest : public ::testing::Test { void testCalcMaxStep() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); n_active = 2; math_t A_host = 3.6534305290498055; math_t ws_host[2] = {0.25662594, -0.01708941}; @@ -470,7 +470,7 @@ class LarsTestFitPredict : public ::testing::Test { void testFitGram() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int max_iter = 10; rapids_logger::level_enum verbosity = rapids_logger::level_enum::off; int n_active; @@ -501,7 +501,7 @@ class LarsTestFitPredict : public ::testing::Test { void testFitX() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int max_iter = 10; rapids_logger::level_enum verbosity = rapids_logger::level_enum::off; int n_active; @@ -532,7 +532,7 @@ class LarsTestFitPredict : public ::testing::Test { void testPredictV1() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int ld_X = n_rows; int n_active = n_cols; raft::update_device(beta.data(), beta_exp, n_active, stream); @@ -555,7 +555,7 @@ class LarsTestFitPredict : public ::testing::Test { void testPredictV2() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int ld_X = n_rows; int n_active = n_cols; @@ -582,7 +582,7 @@ class LarsTestFitPredict : public ::testing::Test { void testFitLarge() { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); int n_rows = 65536; int n_cols = 10; int max_iter = n_cols; diff --git a/cpp/tests/sg/ols.cu b/cpp/tests/sg/ols.cu index 1b1b159dbd..2967158b03 100644 --- a/cpp/tests/sg/ols.cu +++ b/cpp/tests/sg/ols.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -12,6 +12,8 @@ #include #include +#include + #include #include @@ -26,14 +28,16 @@ raft::handle_t create_handle(hconf type) { switch (type) { case hconf::LEGACY_ONE: - return raft::handle_t(rmm::cuda_stream_legacy, std::make_shared(1)); + return raft::handle_t(cuda::stream_ref{cudaStreamLegacy}, + std::make_shared(1)); case hconf::LEGACY_TWO: - return raft::handle_t(rmm::cuda_stream_legacy, std::make_shared(2)); + return raft::handle_t(cuda::stream_ref{cudaStreamLegacy}, + std::make_shared(2)); case hconf::NON_BLOCKING_ONE: - return raft::handle_t(rmm::cuda_stream_per_thread, + return raft::handle_t(cuda::stream_ref{cudaStreamPerThread}, std::make_shared(1)); case hconf::NON_BLOCKING_TWO: - return raft::handle_t(rmm::cuda_stream_per_thread, + return raft::handle_t(cuda::stream_ref{cudaStreamPerThread}, std::make_shared(2)); case hconf::SINGLE: default: return raft::handle_t(); @@ -56,7 +60,7 @@ class OlsTest : public ::testing::TestWithParam> { OlsTest() : params(::testing::TestWithParam>::GetParam()), handle(create_handle(params.hc)), - stream(handle.get_stream()), + stream(handle.get_stream().get()), coef(params.n_col, stream), coef2(params.n_col, stream), coef_ref(params.n_col, stream), diff --git a/cpp/tests/sg/pca_test.cu b/cpp/tests/sg/pca_test.cu index f1e2189ba0..b954947b18 100644 --- a/cpp/tests/sg/pca_test.cu +++ b/cpp/tests/sg/pca_test.cu @@ -42,7 +42,7 @@ class PcaTest : public ::testing::TestWithParam> { public: PcaTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), explained_vars(params.n_col, stream), explained_vars_ref(params.n_col, stream), components(params.n_col * params.n_col, stream), @@ -193,7 +193,7 @@ TEST_P(PcaTestValF, Result) explained_vars_ref.data(), params.n_col, MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestValD; @@ -203,7 +203,7 @@ TEST_P(PcaTestValD, Result) explained_vars_ref.data(), params.n_col, MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestLeftVecF; @@ -213,7 +213,7 @@ TEST_P(PcaTestLeftVecF, Result) components_ref.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestLeftVecD; @@ -223,7 +223,7 @@ TEST_P(PcaTestLeftVecD, Result) components_ref.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestTransDataF; @@ -233,7 +233,7 @@ TEST_P(PcaTestTransDataF, Result) trans_data_ref.data(), (params.n_row * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestTransDataD; @@ -243,7 +243,7 @@ TEST_P(PcaTestTransDataD, Result) trans_data_ref.data(), (params.n_row * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestDataVecSmallF; @@ -253,7 +253,7 @@ TEST_P(PcaTestDataVecSmallF, Result) data_back.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestDataVecSmallD; @@ -263,7 +263,7 @@ TEST_P(PcaTestDataVecSmallD, Result) data_back.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } // FIXME: These tests are disabled due to driver 418+ making them fail: @@ -275,7 +275,7 @@ TEST_P(PcaTestDataVecF, Result) data2_back.data(), (params.n_col2 * params.n_col2), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef PcaTest PcaTestDataVecD; @@ -285,7 +285,7 @@ TEST_P(PcaTestDataVecD, Result) data2_back.data(), (params.n_col2 * params.n_col2), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } INSTANTIATE_TEST_CASE_P(PcaTests, PcaTestValF, ::testing::ValuesIn(inputsf2)); diff --git a/cpp/tests/sg/quasi_newton.cu b/cpp/tests/sg/quasi_newton.cu index 8de627d8d9..91d9bd663a 100644 --- a/cpp/tests/sg/quasi_newton.cu +++ b/cpp/tests/sg/quasi_newton.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -41,7 +41,7 @@ struct QuasiNewtonTest : ::testing::Test { QuasiNewtonTest() : handle(cuml_handle) {} void SetUp() { - stream = cuml_handle.get_stream(); + stream = cuml_handle.get_stream().get(); Xdev.reset(new SimpleMatOwning(N, D, stream, ROW_MAJOR)); raft::update_device(Xdev->data, &X[0][0], Xdev->len, stream); diff --git a/cpp/tests/sg/rf_test.cu b/cpp/tests/sg/rf_test.cu index b3af083742..610cb4dac4 100644 --- a/cpp/tests/sg/rf_test.cu +++ b/cpp/tests/sg/rf_test.cu @@ -20,6 +20,7 @@ #include #include #include +#include #include #include #include @@ -197,9 +198,9 @@ void testBinReductionRoundTrip(std::vector const& input) rmm::device_uvector d_output(input.size(), stream); raft::update_device(d_input.data(), input.data(), input.size(), stream); - DT::packHistograms(d_input.data(), d_packed.data(), input.size(), stream); + DT::packHistograms(d_input.data(), d_packed.data(), input.size(), stream.get()); RAFT_CUDA_TRY(cudaPeekAtLastError()); - DT::unpackHistograms(d_packed.data(), d_output.data(), input.size(), stream); + DT::unpackHistograms(d_packed.data(), d_output.data(), input.size(), stream.get()); RAFT_CUDA_TRY(cudaPeekAtLastError()); std::vector output(input.size()); @@ -254,13 +255,14 @@ std::shared_ptr> nvForestPredict( TreeliteModelHandle model; build_treelite_forest(&model, forest, params.n_cols); - auto nvforest_model = nvforest::import_from_treelite_handle(model, - nvforest::tree_layout::breadth_first, - 128, - std::is_same_v, - nvforest::device_type::gpu, - handle.get_device(), - handle.get_next_usable_stream()); + auto nvforest_model = + nvforest::import_from_treelite_handle(model, + nvforest::tree_layout::breadth_first, + 128, + std::is_same_v, + nvforest::device_type::gpu, + handle.get_device(), + handle.get_next_usable_stream().get()); handle.sync_stream(); handle.sync_stream_pool(); delete static_cast(model); @@ -325,13 +327,14 @@ auto nvForestPredictProba(const raft::handle_t& handle, TreeliteModelHandle model; build_treelite_forest(&model, forest, params.n_cols); - auto nvforest_model = nvforest::import_from_treelite_handle(model, - nvforest::tree_layout::breadth_first, - 128, - std::is_same_v, - nvforest::device_type::gpu, - handle.get_device(), - handle.get_next_usable_stream()); + auto nvforest_model = + nvforest::import_from_treelite_handle(model, + nvforest::tree_layout::breadth_first, + 128, + std::is_same_v, + nvforest::device_type::gpu, + handle.get_device(), + handle.get_next_usable_stream().get()); handle.sync_stream(); handle.sync_stream_pool(); delete static_cast(model); @@ -458,7 +461,7 @@ class RfSpecialisedTest { RfSpecialisedTest(RfTestParams params) : params(params) { auto stream_pool = std::make_shared(params.n_streams); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); X.resize(params.n_rows * params.n_cols); X_transpose.resize(params.n_rows * params.n_cols); y.resize(params.n_rows); @@ -530,7 +533,7 @@ class RfSpecialisedTest { // accuracy is not guaranteed to improve with bootstrapping if (params.bootstrap) { return; } auto stream_pool = std::make_shared(params.n_streams); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); RfTestParams alt_params = params; alt_params.max_depth--; auto [alt_forest, alt_predictions, alt_metrics] = TrainScore(handle, @@ -592,7 +595,7 @@ class RfSpecialisedTest { // Repeat training auto stream_pool = std::make_shared(params.n_streams); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); auto [alt_forest, alt_predictions, alt_metrics] = TrainScore(handle, params, TrainingInputPtr(), @@ -695,7 +698,7 @@ class RfSpecialisedTest { return; } else { auto stream_pool = std::make_shared(params.n_streams); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); auto nvforest_pred = nvForestPredict(handle, params, X_transpose.data().get(), forest.get()); thrust::host_vector h_nvforest_pred(*nvforest_pred); @@ -916,7 +919,7 @@ TEST(RfTests, IntegerOverflow) auto forest = std::make_shared>(); auto forest_ptr = forest.get(); auto stream_pool = std::make_shared(4); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); RF_params rf_params = set_rf_params(3, 100, 1.0, 256, 1, 2, 0.0, false, 1, 1.0, 0, CRITERION::MSE, 4, 128); fit(handle, forest_ptr, X.data().get(), m, n, y.data().get(), rf_params); @@ -929,13 +932,14 @@ TEST(RfTests, IntegerOverflow) TreeliteModelHandle model; build_treelite_forest(&model, forest_ptr, n); - auto nvforest_model = nvforest::import_from_treelite_handle(model, - nvforest::tree_layout::breadth_first, - 128, - false, - nvforest::device_type::gpu, - handle.get_device(), - handle.get_next_usable_stream()); + auto nvforest_model = + nvforest::import_from_treelite_handle(model, + nvforest::tree_layout::breadth_first, + 128, + false, + nvforest::device_type::gpu, + handle.get_device(), + handle.get_next_usable_stream().get()); handle.sync_stream(); handle.sync_stream_pool(); delete static_cast(model); @@ -959,7 +963,7 @@ TEST(RfTests, EmptyGlobalRowsRejected) auto forest = std::make_shared>(); auto forest_ptr = forest.get(); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); RF_params rf_params = set_rf_params(3, 100, 1.0, 16, 1, 2, 0.0, false, 1, 1.0, 0, CRITERION::MSE, 1, 128); @@ -975,7 +979,7 @@ TEST(RfTests, HighClassCountSplitHistogramFallsBackToGlobalMemory) constexpr int max_n_bins = 256; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector X(n_rows * n_cols); thrust::device_vector y(n_rows); @@ -1011,12 +1015,12 @@ TEST(RfTests, InvalidSampleWeightThrows) constexpr std::size_t n_cols = 2; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector X(n_rows * n_cols); thrust::device_vector y(n_rows); thrust::device_vector sample_weight(n_rows, 1.0); raft::random::Rng r(8); - r.normal(X.data().get(), X.size(), 0.0f, 1.0f, handle.get_stream()); + r.normal(X.data().get(), X.size(), 0.0f, 1.0f, handle.get_stream().get()); thrust::host_vector h_y(n_rows); for (std::size_t i = 0; i < n_rows; ++i) { h_y[i] = i % 2; @@ -1027,8 +1031,10 @@ TEST(RfTests, InvalidSampleWeightThrows) set_rf_params(3, 100, 1.0, 8, 1, 2, 0.0, false, 1, 1.0, 0, CRITERION::GINI, 1, 128); auto expect_invalid_weight_throws = [&](double invalid_weight) { - thrust::fill( - thrust::cuda::par.on(handle.get_stream()), sample_weight.begin(), sample_weight.end(), 1.0); + thrust::fill(thrust::cuda::par.on(handle.get_stream().get()), + sample_weight.begin(), + sample_weight.end(), + 1.0); sample_weight[0] = invalid_weight; auto forest = std::make_shared>(); auto forest_ptr = forest.get(); @@ -1049,8 +1055,10 @@ TEST(RfTests, InvalidSampleWeightThrows) expect_invalid_weight_throws(-1.0); expect_invalid_weight_throws(std::numeric_limits::quiet_NaN()); - thrust::fill( - thrust::cuda::par.on(handle.get_stream()), sample_weight.begin(), sample_weight.end(), 0.0); + thrust::fill(thrust::cuda::par.on(handle.get_stream().get()), + sample_weight.begin(), + sample_weight.end(), + 0.0); auto forest = std::make_shared>(); auto forest_ptr = forest.get(); EXPECT_THROW(fit(handle, @@ -1075,13 +1083,13 @@ TEST(RfTests, WeightedBootstrapSamplesOnlyPositiveWeightRows) constexpr int n_zero_weight_rows = 16; auto stream_pool = std::make_shared(2); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector X(n_rows * n_cols); thrust::device_vector y(n_rows); thrust::device_vector sample_weight(n_rows); raft::random::Rng r(8); - r.normal(X.data().get(), X.size(), 0.0f, 1.0f, handle.get_stream()); + r.normal(X.data().get(), X.size(), 0.0f, 1.0f, handle.get_stream().get()); thrust::host_vector h_y(n_rows); thrust::host_vector h_sample_weight(n_rows); @@ -1152,7 +1160,7 @@ class RFQuantileTest : public ::testing::TestWithParam { raft::random::Rng r(8); r.normal(data.data().get(), data.size(), T(0.0), T(2.0), nullptr); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); // computing the quantiles auto quantile_result = @@ -1185,7 +1193,7 @@ class RFQuantileVariableBinsTest : public ::testing::TestWithParam(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector data(params.n_rows); // n_uniques guaranteed to be non-zero and smaller than `max_n_bins` @@ -1254,7 +1262,7 @@ class RFSampledQuantileExactFallbackTest : public ::testing::TestWithParam::GetParam(); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector data(params.n_rows); thrust::sequence(data.begin(), data.end(), T(0)); @@ -1291,7 +1299,7 @@ class RFSampledQuantileDeterminismTest : public ::testing::TestWithParam::GetParam(); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector data(params.n_rows); raft::random::Rng r(params.seed); r.normal(data.data().get(), data.size(), T(0.0), T(2.0), nullptr); @@ -1359,7 +1367,7 @@ TEST(RFEquivalentSplitRangeTest, ClassificationChoosesUpperMiddleBin) constexpr std::int64_t n_bins = 6; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); std::vector h_hist = { {2}, @@ -1383,19 +1391,22 @@ TEST(RFEquivalentSplitRangeTest, ClassificationChoosesUpperMiddleBin) thrust::device_vector mutex(1); DT::ClassificationObjectiveFunction objective(2, 1, CRITERION::GINI); - objectiveGainKernel<<<1, 32, 0, handle.get_stream()>>>(hist.data().get(), - quantiles.data().get(), - split.data().get(), - mutex.data().get(), - objective, - std::int64_t{0}, - len, - n_bins); + objectiveGainKernel<<<1, 32, 0, handle.get_stream().get()>>>(hist.data().get(), + quantiles.data().get(), + split.data().get(), + mutex.data().get(), + objective, + std::int64_t{0}, + len, + n_bins); RAFT_CUDA_TRY(cudaGetLastError()); DT::Split h_split; - RAFT_CUDA_TRY(cudaMemcpyAsync( - &h_split, split.data().get(), sizeof(h_split), cudaMemcpyDeviceToHost, handle.get_stream())); + RAFT_CUDA_TRY(cudaMemcpyAsync(&h_split, + split.data().get(), + sizeof(h_split), + cudaMemcpyDeviceToHost, + handle.get_stream().get())); handle.sync_stream(); EXPECT_EQ(h_split.global_nLeft, 4); @@ -1412,7 +1423,7 @@ TEST(RFEquivalentSplitRangeTest, RegressionChoosesUpperMiddleBin) constexpr std::int64_t n_bins = 6; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); std::vector h_hist = { {0.0, 2}, @@ -1430,19 +1441,22 @@ TEST(RFEquivalentSplitRangeTest, RegressionChoosesUpperMiddleBin) thrust::device_vector mutex(1); DT::RegressionObjectiveFunction objective(1, 1, CRITERION::MSE); - objectiveGainKernel<<<1, 32, 0, handle.get_stream()>>>(hist.data().get(), - quantiles.data().get(), - split.data().get(), - mutex.data().get(), - objective, - std::int64_t{0}, - len, - n_bins); + objectiveGainKernel<<<1, 32, 0, handle.get_stream().get()>>>(hist.data().get(), + quantiles.data().get(), + split.data().get(), + mutex.data().get(), + objective, + std::int64_t{0}, + len, + n_bins); RAFT_CUDA_TRY(cudaGetLastError()); DT::Split h_split; - RAFT_CUDA_TRY(cudaMemcpyAsync( - &h_split, split.data().get(), sizeof(h_split), cudaMemcpyDeviceToHost, handle.get_stream())); + RAFT_CUDA_TRY(cudaMemcpyAsync(&h_split, + split.data().get(), + sizeof(h_split), + cudaMemcpyDeviceToHost, + handle.get_stream().get())); handle.sync_stream(); EXPECT_EQ(h_split.global_nLeft, 4); @@ -1460,7 +1474,7 @@ class RFSampledQuantileRankErrorTest : public ::testing::TestWithParam::GetParam(); auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); thrust::device_vector data(params.n_rows); thrust::sequence(data.begin(), data.end(), T(0)); @@ -1596,7 +1610,7 @@ TEST(RfTest, TextDump) thrust::device_vector y = y_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); auto forest_ptr = forest.get(); fit(handle, forest_ptr, X.data().get(), y.size(), 1, y.data().get(), 2, rf_params); @@ -1635,7 +1649,7 @@ TEST(RfTest, EquivalentSplitRangePersistsThroughBuilder) thrust::device_vector y = y_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); auto forest_ptr = forest.get(); fit(handle, forest_ptr, X.data().get(), y.size(), 2, y.data().get(), 2, rf_params); @@ -1674,7 +1688,7 @@ TEST(RfWeightedTest, ClassificationRootLeafUsesWeights) std::vector weight_host = {100.0f, 1.0f, 1.0f}; thrust::device_vector weights = weight_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); fit(handle, forest.get(), @@ -1710,7 +1724,7 @@ TEST(RfWeightedTest, RegressionRootLeafUsesWeights) std::vector weight_host = {1.0f, 0.0f, 3.0f}; thrust::device_vector weights = weight_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); fit(handle, forest.get(), @@ -1744,7 +1758,7 @@ TEST(RfWeightedTest, MinSamplesLeafUsesCountsNotWeights) std::vector weight_host = {0.1f, 0.1f, 100.0f, 100.0f}; thrust::device_vector weights = weight_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); fit(handle, forest.get(), @@ -1784,7 +1798,7 @@ TEST(RfWeightedTest, ZeroWeightSamplesDoNotCreatePositiveWeightSplit) std::vector weight_host = {0.0f, 0.0f, 1.0f, 1.0f}; thrust::device_vector weights = weight_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); fit(handle, forest.get(), @@ -1817,7 +1831,7 @@ TEST(RfWeightedTest, BootstrapDuplicatesContributePerOccurrence) std::vector weight_host = {1.0f, 2.0f, 5.0f}; thrust::device_vector weights = weight_host; auto stream_pool = std::make_shared(1); - raft::handle_t handle(rmm::cuda_stream_per_thread, stream_pool); + raft::handle_t handle(cuda::stream_ref{cudaStreamPerThread}, stream_pool); constexpr int n_rows = 3; bool found_duplicate = false; @@ -2644,7 +2658,7 @@ class FeatureSamplingBiasTest : public ::testing::TestWithParam::GetParam(); stream_pool = std::make_shared(1); - handle.reset(new raft::handle_t(rmm::cuda_stream_per_thread, stream_pool)); + handle.reset(new raft::handle_t(cuda::stream_ref{cudaStreamPerThread}, stream_pool)); } void TearDown() override diff --git a/cpp/tests/sg/ridge.cu b/cpp/tests/sg/ridge.cu index 6f9509ae6e..f0987ec9ef 100644 --- a/cpp/tests/sg/ridge.cu +++ b/cpp/tests/sg/ridge.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -32,7 +32,7 @@ class RidgeTest : public ::testing::TestWithParam> { public: RidgeTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), coef(params.n_col, stream), coef2(params.n_col, stream), coef_ref(params.n_col, stream), diff --git a/cpp/tests/sg/sgd.cu b/cpp/tests/sg/sgd.cu index c106dc1fe8..befa46edcd 100644 --- a/cpp/tests/sg/sgd.cu +++ b/cpp/tests/sg/sgd.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -32,7 +32,7 @@ class SgdTest : public ::testing::TestWithParam> { public: SgdTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), coef(params.n_col, stream), coef_ref(params.n_col, stream), coef2(params.n_col, stream), diff --git a/cpp/tests/sg/shap_kernel.cu b/cpp/tests/sg/shap_kernel.cu index c95c554d26..37a03ab870 100644 --- a/cpp/tests/sg/shap_kernel.cu +++ b/cpp/tests/sg/shap_kernel.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -42,7 +42,7 @@ class MakeKSHAPDatasetTest : public ::testing::TestWithParam::GetParam(); - stream = handle.get_stream(); + stream = handle.get_stream().get(); int i, j; nrows_X = params.nrows_exact + params.nrows_sampled; diff --git a/cpp/tests/sg/svc_test.cu b/cpp/tests/sg/svc_test.cu index 2d1e4ff5f6..25a8402b49 100644 --- a/cpp/tests/sg/svc_test.cu +++ b/cpp/tests/sg/svc_test.cu @@ -65,7 +65,7 @@ template class WorkingSetTest : public ::testing::Test { public: WorkingSetTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), f_dev(10, stream), y_dev(10, stream), C_dev(10, stream), @@ -105,7 +105,7 @@ TYPED_TEST_CASE(WorkingSetTest, FloatTypes); TYPED_TEST(WorkingSetTest, Init) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); this->ws = new WorkingSet(this->handle, stream, 10); EXPECT_EQ(this->ws->GetSize(), 10); delete this->ws; @@ -117,7 +117,7 @@ TYPED_TEST(WorkingSetTest, Init) TYPED_TEST(WorkingSetTest, Select) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); this->ws = new WorkingSet(this->handle, stream, 10, 4); EXPECT_EQ(this->ws->GetSize(), 4); this->ws->SimpleSelect( @@ -167,7 +167,7 @@ template class KernelCacheTest : public ::testing::Test { public: KernelCacheTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), n_rows(4), n_cols(2), n_ws(3), @@ -221,7 +221,7 @@ class KernelCacheTest : public ::testing::Test { void check(math_t* kernel_data, int* nz_da_idx, int nnz_da, int batch_size, int offset) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); std::vector ws_idx_h(nnz_da); raft::update_host(ws_idx_h.data(), nz_da_idx, nnz_da, stream); handle.sync_stream(stream); @@ -265,7 +265,7 @@ TYPED_TEST_CASE_P(KernelCacheTest); TYPED_TEST_P(KernelCacheTest, EvalTest) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); std::vector param_vec{{KernelType::LINEAR, 3, 1, 0}, {KernelType::POLYNOMIAL, 2, 1.3, 1}, {KernelType::TANH, 2, 0.5, 2.4}, @@ -496,7 +496,7 @@ INSTANTIATE_TYPED_TEST_CASE_P(My, KernelCacheTest, FloatTypes); template class GetResultsTest : public ::testing::Test { public: - GetResultsTest() : stream(handle.get_stream()) {} + GetResultsTest() : stream(handle.get_stream().get()) {} protected: void FreeDenseSupport() @@ -509,7 +509,7 @@ class GetResultsTest : public ::testing::Test { void TestResults() { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); rmm::device_uvector x_dev(n_rows * n_cols, stream); raft::update_device(x_dev.data(), x_host, n_rows * n_cols, stream); rmm::device_uvector f_dev(n_rows, stream); @@ -597,7 +597,7 @@ template class SmoUpdateTest : public ::testing::Test { public: SmoUpdateTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), n_rows(6), n_ws(2), f_dev(n_rows, stream), @@ -639,7 +639,7 @@ template class SmoBlockSolverTest : public ::testing::Test { public: SmoBlockSolverTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), n_rows(4), n_cols(2), n_ws(4), @@ -870,7 +870,7 @@ template class SmoSolverTest : public ::testing::Test { public: SmoSolverTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), x_dev(n_rows * n_cols, stream), x_dev_indptr(n_rows + 1, stream), x_dev_indices(n_nnz, stream), @@ -973,7 +973,7 @@ class SmoSolverTest : public ::testing::Test { void svrBlockSolveTest() { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); int n_ws = 4; int n_rows = 2; // int n_cols = 1; @@ -1080,7 +1080,7 @@ std::ostream& operator<<(std::ostream& os, const smoInput& b) TYPED_TEST(SmoSolverTest, SmoSolveTest) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); std::vector, smoOutput>> data{ {smoInput{1, 0.001, KernelParams{KernelType::LINEAR, 3, 1, 0}, 100, 1}, smoOutput{4, // n_sv @@ -1166,7 +1166,7 @@ TYPED_TEST(SmoSolverTest, SmoSolveTest) TYPED_TEST(SmoSolverTest, SvcTest) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); std::vector, smoOutput2>> data{ {svcInput{1, 0.001, @@ -1313,7 +1313,7 @@ void make_blobs(const raft::handle_t& handle, size_t free1, total; RAFT_CUDA_TRY(cudaMemGetInfo(&free1, &total)); { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); rmm::device_uvector x_float(n_rows * n_cols, stream); rmm::device_uvector y_int(n_rows, stream); @@ -1358,7 +1358,7 @@ struct is_same_functor { TYPED_TEST(SmoSolverTest, BlobPredict) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); // Pair.second is the expected accuracy. It might change if the Rng changes. std::vector> data{ {blobInput{1, 0.001, KernelParams{KernelType::LINEAR, 3, 1, 0}, 200, 10}, 98}, @@ -1414,7 +1414,7 @@ TYPED_TEST(SmoSolverTest, MemoryLeak) { GTEST_SKIP(); // Skip the tests in CI for release 24.02 // https://github.com/NVIDIA/cuml/issues/5763 - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); // We measure that we have the same amount of free memory available on the GPU // before and after we call SVM. This can help catch memory leaks, but it is // not 100% sure. Small allocations might be pooled together by cudaMalloc, @@ -1476,7 +1476,7 @@ TYPED_TEST(SmoSolverTest, MemoryLeak) TYPED_TEST(SmoSolverTest, DISABLED_MillionRows) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); if (sizeof(TypeParam) == 8) { GTEST_SKIP(); // Skip the test for double input } else { @@ -1525,7 +1525,7 @@ template void initializeTestMatrix( const raft::handle_t& handle, math_t* dense_matrix, int n_rows, int n_cols, math_t* y) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); assert(n_cols % n_rows * n_rows % n_cols == 0); /* @@ -1580,7 +1580,7 @@ void initializeTestMatrix(const raft::handle_t& handle, int n_cols, math_t* y) { - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); assert(n_cols % n_rows * n_rows % n_cols == 0); /* @@ -1707,7 +1707,7 @@ TYPED_TEST(SmoSolverTest, DenseBatching) TYPED_TEST(SmoSolverTest, SparseBatching) { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); if (sizeof(TypeParam) == 8) { GTEST_SKIP(); // Skip the test for double input } else { @@ -1855,7 +1855,7 @@ template class SvrTest : public ::testing::Test { public: SvrTest() - : stream(handle.get_stream()), + : stream(handle.get_stream().get()), x_dev(n_rows * n_cols, stream), y_dev(n_rows, stream), C_dev(2 * n_rows, stream), @@ -1884,7 +1884,7 @@ class SvrTest : public ::testing::Test { public: void TestSvrInit() { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); SvmParameter param = getDefaultSvmParameter(); param.svmType = EPSILON_SVR; SmoSolver smo(handle, param, cuvs::distance::kernels::KernelType::LINEAR, nullptr); @@ -1965,7 +1965,7 @@ class SvrTest : public ::testing::Test { void TestSvrFitPredict() { - auto stream = this->handle.get_stream(); + auto stream = this->handle.get_stream().get(); std::vector, smoOutput2>> data{ {SvrInput{ SvmParameter{1, 0, 1, -1, 10, 1e-3, rapids_logger::level_enum::info, 0.1, EPSILON_SVR}, diff --git a/cpp/tests/sg/trustworthiness_test.cu b/cpp/tests/sg/trustworthiness_test.cu index 593f3eb98e..22f3556daa 100644 --- a/cpp/tests/sg/trustworthiness_test.cu +++ b/cpp/tests/sg/trustworthiness_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -300,7 +300,7 @@ class TrustworthinessScoreTest : public ::testing::Test { -0.30633628}; raft::handle_t h; - cudaStream_t stream = h.get_stream(); + cudaStream_t stream = h.get_stream().get(); rmm::device_uvector d_X(X.size(), stream); rmm::device_uvector d_X_embedded(X_embedded.size(), stream); diff --git a/cpp/tests/sg/tsne_test.cu b/cpp/tests/sg/tsne_test.cu index 74e287f7ef..ea4a81fac3 100644 --- a/cpp/tests/sg/tsne_test.cu +++ b/cpp/tests/sg/tsne_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2019-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2019-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -102,7 +102,7 @@ class TSNETest : public ::testing::TestWithParam { TSNEResults runTest(TSNE_ALGORITHM algo, bool knn = false) { raft::handle_t handle; - auto stream = handle.get_stream(); + auto stream = handle.get_stream().get(); TSNEResults results; auto DEFAULT_DISTANCE_METRIC = ML::distance::DistanceType::L2SqrtExpanded; diff --git a/cpp/tests/sg/tsvd_test.cu b/cpp/tests/sg/tsvd_test.cu index 197053ad7c..864cc53409 100644 --- a/cpp/tests/sg/tsvd_test.cu +++ b/cpp/tests/sg/tsvd_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2018-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -40,7 +40,7 @@ class TsvdTest : public ::testing::TestWithParam> { public: TsvdTest() : params(::testing::TestWithParam>::GetParam()), - stream(handle.get_stream()), + stream(handle.get_stream().get()), components(0, stream), components_ref(0, stream), data2(0, stream), @@ -160,7 +160,7 @@ TEST_P(TsvdTestLeftVecF, Result) components_ref.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef TsvdTest TsvdTestLeftVecD; @@ -170,7 +170,7 @@ TEST_P(TsvdTestLeftVecD, Result) components_ref.data(), (params.n_col * params.n_col), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef TsvdTest TsvdTestDataVecF; @@ -180,7 +180,7 @@ TEST_P(TsvdTestDataVecF, Result) data2_back.data(), (params.n_col2 * params.n_col2), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } typedef TsvdTest TsvdTestDataVecD; @@ -190,7 +190,7 @@ TEST_P(TsvdTestDataVecD, Result) data2_back.data(), (params.n_col2 * params.n_col2), MLCommon::CompareApprox(params.tolerance), - handle.get_stream())); + handle.get_stream().get())); } INSTANTIATE_TEST_CASE_P(TsvdTests, TsvdTestLeftVecF, ::testing::ValuesIn(inputsf2)); diff --git a/cpp/tests/sg/umap_parametrizable_test.cu b/cpp/tests/sg/umap_parametrizable_test.cu index ece408fbbd..1aae15c992 100644 --- a/cpp/tests/sg/umap_parametrizable_test.cu +++ b/cpp/tests/sg/umap_parametrizable_test.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2020-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2020-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -105,7 +105,7 @@ class UMAPParametrizableTest : public ::testing::Test { TestParams& test_params, UMAPParams& umap_params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int& n_samples = test_params.n_samples; int& n_features = test_params.n_features; @@ -219,7 +219,7 @@ class UMAPParametrizableTest : public ::testing::Test { TestParams& test_params, UMAPParams& umap_params) { - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int& n_samples = test_params.n_samples; int& n_features = test_params.n_features; @@ -253,7 +253,7 @@ class UMAPParametrizableTest : public ::testing::Test { << test_params.min_trustworthiness << "]" << std::endl; raft::handle_t handle; - cudaStream_t stream = handle.get_stream(); + cudaStream_t stream = handle.get_stream().get(); int& n_samples = test_params.n_samples; int& n_features = test_params.n_features;