From 5355fda122388765d309dc4c85ebc564af9b9233 Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Tue, 1 Sep 2026 05:52:50 -0500 Subject: [PATCH 1/5] Migrate stream view APIs to cuda::stream_ref --- c/src/core/c_api.cpp | 2 +- cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h | 4 +- .../cuvs_internal/neighbors/refine_helper.cuh | 2 +- cpp/src/neighbors/detail/ann_utils.cuh | 29 +++++------ .../detail/cagra/compute_distance.hpp | 13 +++-- .../cagra/compute_distance_standard-impl.cuh | 2 +- .../cagra/compute_distance_vpq-impl.cuh | 3 +- .../neighbors/detail/cagra/search_plan.cuh | 8 +-- cpp/src/neighbors/ivf_common.cu | 4 +- cpp/src/neighbors/ivf_common.cuh | 6 +-- ...vf_flat_interleaved_scan_explicit_inst.cuh | 2 +- .../ivf_flat_interleaved_scan_ext.cuh | 4 +- .../ivf_flat_interleaved_scan_jit.cuh | 4 +- .../ivf_pq_compute_similarity_run_inst.cu.in | 2 +- .../detail/ivf_pq_contiguous_list_data.cu | 4 +- .../ivf_pq/detail/ivf_pq_list_data.cu | 4 +- cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh | 2 +- .../ivf_pq/ivf_pq_compute_similarity.cuh | 6 +-- .../ivf_pq/ivf_pq_compute_similarity_impl.cuh | 4 +- .../ivf_pq/ivf_pq_contiguous_list_data.cuh | 4 +- .../ivf_pq_contiguous_list_data_impl.cuh | 4 +- cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp | 4 +- .../ivf_pq/ivf_pq_list_data_impl.cuh | 4 +- .../ivf_rabitq/gpu_index/initializer_gpu.cuh | 2 +- .../neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu | 2 +- .../ivf_rabitq/gpu_index/ivf_gpu.cuh | 2 +- .../ivf_rabitq/gpu_index/quantizer_gpu.cuh | 2 +- .../ivf_rabitq/gpu_index/searcher_gpu.cu | 2 +- .../ivf_rabitq/gpu_index/searcher_gpu.cuh | 4 +- cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh | 4 +- cpp/tests/neighbors/all_neighbors.cuh | 2 +- cpp/tests/neighbors/ann_brute_force.cuh | 2 +- cpp/tests/neighbors/ann_cagra.cuh | 12 ++--- cpp/tests/neighbors/ann_hnsw_ace.cuh | 2 +- cpp/tests/neighbors/ann_ivf_flat.cuh | 2 +- cpp/tests/neighbors/ann_ivf_pq.cuh | 4 +- cpp/tests/neighbors/ann_ivf_rabitq.cuh | 2 +- cpp/tests/neighbors/ann_ivf_sq.cuh | 2 +- cpp/tests/neighbors/ann_nn_descent.cuh | 6 +-- cpp/tests/neighbors/ann_scann.cuh | 2 +- cpp/tests/neighbors/ann_vamana.cuh | 2 +- cpp/tests/neighbors/distance_nn.cu | 2 +- cpp/tests/neighbors/hnsw.cu | 2 +- cpp/tests/neighbors/refine.cu | 2 +- cpp/tests/neighbors/refine_helper.cuh | 2 +- cpp/tests/neighbors/tiered_index.cu | 2 +- ...pes-copy-serialization-and-utility-apis.md | 16 +++--- ...pp-api-common-types-execution-resources.md | 36 ++++++------- fern/pages/other/resources.md | 2 +- fern/scripts/generate_api_reference.py | 52 +++++++++---------- 50 files changed, 145 insertions(+), 148 deletions(-) diff --git a/c/src/core/c_api.cpp b/c/src/core/c_api.cpp index 5d431afcf8..ca4e7fcd90 100644 --- a/c/src/core/c_api.cpp +++ b/c/src/core/c_api.cpp @@ -136,7 +136,7 @@ extern "C" cuvsError_t cuvsStreamSet(cuvsResources_t res, cudaStream_t stream) { return cuvs::core::translate_exceptions([=] { auto res_ptr = reinterpret_cast(res); - raft::resource::set_cuda_stream(*res_ptr, static_cast(stream)); + raft::resource::set_cuda_stream(*res_ptr, static_cast(stream)); }); } diff --git a/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h b/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h index 1a276e8cc8..64e43412dd 100644 --- a/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h +++ b/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h @@ -122,8 +122,8 @@ class configured_raft_resources { */ explicit configured_raft_resources(const std::shared_ptr& shared_res) : shared_res_{shared_res}, - res_{std::make_unique( - rmm::cuda_stream_view(get_stream_from_global_pool()))} + res_{ + std::make_unique(cuda::stream_ref(get_stream_from_global_pool()))} { raft::resource::set_large_workspace_resource( *res_, raft::mr::device_resource{shared_res_->get_large_memory_resource()}); diff --git a/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh b/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh index d5c43aaa34..3aa7b78bb9 100644 --- a/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh +++ b/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh @@ -127,7 +127,7 @@ class RefineHelper { public: RefineInputs p; const raft::resources& handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; raft::device_matrix dataset; raft::device_matrix queries; diff --git a/cpp/src/neighbors/detail/ann_utils.cuh b/cpp/src/neighbors/detail/ann_utils.cuh index b3c5b11f27..a803ed3ca5 100644 --- a/cpp/src/neighbors/detail/ann_utils.cuh +++ b/cpp/src/neighbors/detail/ann_utils.cuh @@ -224,7 +224,7 @@ HDI constexpr auto mapping::operator()(const float& x) const -> int8_t * @param[in] n_bytes */ template -inline void memzero(T* ptr, IdxT n_elems, rmm::cuda_stream_view stream) +inline void memzero(T* ptr, IdxT n_elems, cuda::stream_ref stream) { switch (check_pointer_residency(ptr)) { case pointer_residency::host_and_device: @@ -298,7 +298,7 @@ void block_copy(const IdxT* in_offsets, const T* in_data, T* out_data, IdxT n_mult, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { IdxT in_size; update_host(&in_size, in_offsets + n_blocks, 1, stream); @@ -325,7 +325,7 @@ void block_copy(const IdxT* in_offsets, * @param stream */ template -void outer_add(const T* a, IdxT len_a, const T* b, IdxT len_b, T* c, rmm::cuda_stream_view stream) +void outer_add(const T* a, IdxT len_a, const T* b, IdxT len_b, T* c, cuda::stream_ref stream) { dim3 threads(128, 1, 1); dim3 blocks(raft::ceildiv(len_a * len_b, threads.x), 1, 1); @@ -370,7 +370,7 @@ void copy_selected(IdxT n_rows, IdxT ld_src, T* dst, IdxT ld_dst, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { switch (check_pointer_residency(src, dst, row_ids)) { case pointer_residency::host_and_device: @@ -403,8 +403,7 @@ void copy_selected(IdxT n_rows, * the main stream itself is returned with `false`, and the caller should treat prefetch as a * no-op (no overlap is possible on a single stream). */ -inline auto get_prefetch_stream(raft::resources const& res) - -> std::pair +inline auto get_prefetch_stream(raft::resources const& res) -> std::pair { if (res.has_resource_factory(raft::resource::resource_type::CUDA_STREAM_POOL) && raft::resource::get_stream_pool_size(res) >= 1) { @@ -600,7 +599,7 @@ struct batch_load_iterator { batch(raft::resources const& res, MdspanT input_view, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, rmm::device_async_resource_ref mr, bool prefetch, bool initialize, @@ -806,7 +805,7 @@ struct batch_load_iterator { copy_stream_.get())); } - rmm::cuda_stream_view copy_stream_; + cuda::stream_ref copy_stream_; raft::resources const* res_; MdspanT input_view_; element_type* source_; @@ -860,7 +859,7 @@ struct batch_load_iterator { batch_load_iterator(raft::resources const& res, MdspanT input_view, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, rmm::device_async_resource_ref mr, bool prefetch = false, bool initialize = true, @@ -876,7 +875,7 @@ struct batch_load_iterator { batch_load_iterator(raft::resources const& res, MdspanT input_view, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, bool prefetch = false, bool initialize = true, bool host_writeback = false) @@ -1025,7 +1024,7 @@ class batch_load_iterator_dyn { IdxT n_rows, IdxT row_width, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, rmm::device_async_resource_ref mr, bool prefetch = false, bool initialize = true, @@ -1050,7 +1049,7 @@ class batch_load_iterator_dyn { IdxT n_rows, IdxT row_width, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, bool prefetch = false, bool initialize = true, bool host_writeback = false) @@ -1156,7 +1155,7 @@ class batch_load_iterator_dyn { IdxT n_rows, IdxT row_width, size_type batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, rmm::device_async_resource_ref mr, bool prefetch, bool initialize, @@ -1219,7 +1218,7 @@ auto make_batch_load_iterator(raft::resources const& res, detail::type_identity_t n_rows, detail::type_identity_t row_width, size_t batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, rmm::device_async_resource_ref mr, bool prefetch = false, bool initialize = true, @@ -1244,7 +1243,7 @@ auto make_batch_load_iterator(raft::resources const& res, detail::type_identity_t n_rows, detail::type_identity_t row_width, size_t batch_size, - rmm::cuda_stream_view copy_stream, + cuda::stream_ref copy_stream, bool prefetch = false, bool initialize = true, bool host_writeback = false) -> batch_load_iterator_dyn diff --git a/cpp/src/neighbors/detail/cagra/compute_distance.hpp b/cpp/src/neighbors/detail/cagra/compute_distance.hpp index 55fe39d585..f080932cdc 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance.hpp +++ b/cpp/src/neighbors/detail/cagra/compute_distance.hpp @@ -211,9 +211,8 @@ struct dataset_descriptor_host { // Codebook type is determined by DataT for VPQ (always half for now) struct state { - using ready_t = std::tuple; - using init_f = - std::tuple, size_t>; + using ready_t = std::tuple; + using init_f = std::tuple, size_t>; std::mutex mutex; std::atomic ready; // Not sure if std::holds_alternative is thread-safe @@ -235,7 +234,7 @@ struct dataset_descriptor_host { RAFT_CUDA_TRY_NO_THROW(cudaEventDestroy(ready_event)); } - void eval(rmm::cuda_stream_view stream) + void eval(cuda::stream_ref stream) { std::lock_guard lock(mutex); if (std::holds_alternative(value)) { @@ -249,7 +248,7 @@ struct dataset_descriptor_host { } } - auto get(rmm::cuda_stream_view stream) -> dev_descriptor_t* + auto get(cuda::stream_ref stream) -> dev_descriptor_t* { if (!ready.load(std::memory_order_acquire)) { eval(stream); } // value is immutable at this point. @@ -288,12 +287,12 @@ struct dataset_descriptor_host { /** * Return the device pointer, possibly evaluating it in the given thread. */ - [[nodiscard]] auto dev_ptr(rmm::cuda_stream_view stream) const -> const dev_descriptor_t* + [[nodiscard]] auto dev_ptr(cuda::stream_ref stream) const -> const dev_descriptor_t* { return value_->get(stream); } - [[nodiscard]] auto dev_ptr(rmm::cuda_stream_view stream) -> dev_descriptor_t* + [[nodiscard]] auto dev_ptr(cuda::stream_ref stream) -> dev_descriptor_t* { return value_->get(stream); } diff --git a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh index 06cdb375da..59b40167b3 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh +++ b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh @@ -133,7 +133,7 @@ standard_descriptor_spec* dev_ptr, - rmm::cuda_stream_view stream) { + cuda::stream_ref stream) { standard_dataset_descriptor_init_kernel* dev_ptr, - rmm::cuda_stream_view stream) { + [=](dataset_descriptor_base_t* dev_ptr, cuda::stream_ref stream) { vpq_dataset_descriptor_init_kernel struct lightweight_uvector { private: - using raft_res_type = const raft::resources*; - using rmm_res_type = std::tuple; + using raft_res_type = const raft::resources*; + using rmm_res_type = std::tuple; static constexpr size_t kAlign = 256; std::variant res_; @@ -75,14 +75,14 @@ struct lightweight_uvector { size_ = new_size; } - void resize(size_t new_size, rmm::cuda_stream_view stream) + void resize(size_t new_size, cuda::stream_ref stream) { if (new_size == size_) { return; } if (std::holds_alternative(res_)) { auto& h = std::get(res_); res_ = rmm_res_type{raft::resource::get_workspace_resource_ref(*h), stream}; } else { - std::get(std::get(res_)) = stream; + std::get(std::get(res_)) = stream; } resize(new_size); } diff --git a/cpp/src/neighbors/ivf_common.cu b/cpp/src/neighbors/ivf_common.cu index 7551ef1214..ec7da858d8 100644 --- a/cpp/src/neighbors/ivf_common.cu +++ b/cpp/src/neighbors/ivf_common.cu @@ -53,7 +53,7 @@ void calc_chunk_indices::configured::operator()(const uint32_t* cluster_sizes, const uint32_t* clusters_to_probe, uint32_t* chunk_indices, uint32_t* n_samples, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { void* kernel = nullptr; switch (block_dim.x) { @@ -78,7 +78,7 @@ void calc_chunk_indices::configured::operator()(const uint32_t* cluster_sizes, void sort_cluster_sizes_descending(uint32_t* input, uint32_t* output, uint32_t n_lists, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, rmm::device_async_resource_ref tmp_res) { int begin_bit = 0; diff --git a/cpp/src/neighbors/ivf_common.cuh b/cpp/src/neighbors/ivf_common.cuh index fb1513065d..9dd3cd15bb 100644 --- a/cpp/src/neighbors/ivf_common.cuh +++ b/cpp/src/neighbors/ivf_common.cuh @@ -19,7 +19,7 @@ namespace cuvs::neighbors::ivf::detail { void sort_cluster_sizes_descending(uint32_t* input, uint32_t* output, uint32_t n_lists, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, rmm::device_async_resource_ref tmp_res); /** @@ -57,7 +57,7 @@ struct calc_chunk_indices { const uint32_t* clusters_to_probe, uint32_t* chunk_indices, uint32_t* n_samples, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); }; static inline auto configure(uint32_t n_probes, uint32_t n_queries) -> configured @@ -153,7 +153,7 @@ void postprocess_neighbors(IdxT* neighbors_out, // [n_queries, to uint32_t n_queries, uint32_t n_probes, uint32_t topk, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { constexpr int kPNThreads = 256; const int pn_blocks = raft::div_rounding_up_unsafe(n_queries * topk, kPNThreads); diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh index 052b7bfe9a..f5f61bc786 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh @@ -36,7 +36,7 @@ uint32_t* neighbors, \ float* distances, \ uint32_t& grid_dim_x, \ - rmm::cuda_stream_view stream, \ + cuda::stream_ref stream, \ const std::optional& metric_udf); #define COMMA , diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh index 3a782822b4..e5d8da76d6 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh @@ -35,7 +35,7 @@ void ivfflat_interleaved_scan(const index& index, uint32_t* neighbors, float* distances, uint32_t& grid_dim_x, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, const std::optional& metric_udf) RAFT_EXPLICIT; #define CUVS_INST_IVF_FLAT_INTERLEAVED_SCAN(T, IdxT, SampleFilterT) \ @@ -58,7 +58,7 @@ void ivfflat_interleaved_scan(const index& index, uint32_t* neighbors, \ float* distances, \ uint32_t& grid_dim_x, \ - rmm::cuda_stream_view stream, \ + cuda::stream_ref stream, \ const std::optional& metric_udf); CUVS_INST_IVF_FLAT_INTERLEAVED_SCAN(float, int64_t, cuvs::neighbors::filtering::none_sample_filter); diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh index cf4fface36..c9bfc93e7f 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh @@ -151,7 +151,7 @@ void launch_kernel(const index& index, uint32_t* neighbors, float* distances, uint32_t& grid_dim_x, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, const std::optional& metric_udf) { RAFT_EXPECTS(Veclen == index.veclen(), @@ -435,7 +435,7 @@ void ivfflat_interleaved_scan(const index& index, uint32_t* neighbors, float* distances, uint32_t& grid_dim_x, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, const std::optional& metric_udf) { const uint32_t n_probes_clamped = std::min(n_probes, index.n_lists()); diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in index 05892b9333..320f6bd98e 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in @@ -17,7 +17,7 @@ namespace cuvs::neighbors::ivf_pq::detail { template void cuvs::neighbors::ivf_pq::detail::compute_similarity_run( cuvs::neighbors::ivf_pq::detail::selected s, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, uint32_t dim, uint32_t n_probes, uint32_t pq_dim, diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu index 5fac491532..d215c69202 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu @@ -16,7 +16,7 @@ void unpack_contiguous_list_data( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { unpack_contiguous_list_data_impl( codes, list_data, n_rows, pq_dim, offset_or_indices, pq_bits, stream); @@ -31,7 +31,7 @@ void pack_contiguous_list_data( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { pack_contiguous_list_data_impl( list_data, codes, n_rows, pq_dim, offset_or_indices, pq_bits, stream); diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu index 34a7244a67..c439298f04 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu @@ -13,7 +13,7 @@ void unpack_list_data(raft::device_matrix_view list_data, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { unpack_list_data_impl(codes, list_data, offset_or_indices, pq_bits, stream); }; @@ -24,7 +24,7 @@ void pack_list_data(raft::device_mdspan codes, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { pack_list_data_impl(list_data, codes, offset_or_indices, pq_bits, stream); }; diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh index 78177b7acf..187b382b9f 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh @@ -235,7 +235,7 @@ auto calculate_offsets_and_indices(IdxT n_rows, const uint32_t* cluster_sizes, IdxT* cluster_offsets, IdxT* data_indices, - rmm::cuda_stream_view stream) -> uint32_t + cuda::stream_ref stream) -> uint32_t { auto exec_policy = rmm::exec_policy(stream); // Calculate the offsets diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh index c55e3cdbc0..34e9b760eb 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh @@ -14,7 +14,7 @@ #include #include // cuvs::neighbors::ivf_pq::codebook_gen #include // RAFT_WEAK_FUNCTION -#include // rmm::cuda_stream_view +#include // cuda::stream_ref #include // __half @@ -28,7 +28,7 @@ auto RAFT_WEAK_FUNCTION is_local_topk_feasible(uint32_t k, uint32_t n_probes, ui template void compute_similarity_run(selected s, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, uint32_t dim, uint32_t n_probes, uint32_t pq_dim, @@ -138,7 +138,7 @@ auto compute_similarity_select(const cudaDeviceProp& dev_props, \ extern template void cuvs::neighbors::ivf_pq::detail::compute_similarity_run( \ cuvs::neighbors::ivf_pq::detail::selected s, \ - rmm::cuda_stream_view stream, \ + cuda::stream_ref stream, \ uint32_t dim, \ uint32_t n_probes, \ uint32_t pq_dim, \ diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh index 19f6bb6dc4..212e596cae 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh @@ -21,7 +21,7 @@ #include // raft::Pow2 #include // raft::TxN_t -#include // rmm::cuda_stream_view +#include // cuda::stream_ref namespace cuvs::neighbors::ivf_pq::detail { @@ -265,7 +265,7 @@ struct occupancy_t { template void compute_similarity_run(selected s, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, uint32_t dim, uint32_t n_probes, uint32_t pq_dim, diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh index 87c4c4ea79..507528e905 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh @@ -20,7 +20,7 @@ void unpack_contiguous_list_data( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); template void unpack_contiguous_list_data(raft::resources const& res, @@ -53,7 +53,7 @@ void pack_contiguous_list_data( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); template void pack_contiguous_list_data(raft::resources const& res, diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh index a473324b60..369f9ad31c 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh @@ -72,7 +72,7 @@ inline void unpack_contiguous_list_data_impl( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { if (n_rows == 0) { return; } @@ -154,7 +154,7 @@ inline void pack_contiguous_list_data_impl( uint32_t pq_dim, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { if (n_rows == 0) { return; } diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp index ac6bbec662..774a2a4ef5 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp @@ -18,7 +18,7 @@ void unpack_list_data(raft::device_matrix_view list_data, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); void pack_list_data(raft::device_mdspan::list_extents, @@ -26,7 +26,7 @@ void pack_list_data(raft::device_mdspan codes, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); /** Unpack the list data; see the public interface for the api and usage. */ template diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh index aad5f428b1..ab2df79966 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh @@ -68,7 +68,7 @@ inline void unpack_list_data_impl( raft::row_major> list_data, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { auto n_rows = codes.extent(0); if (n_rows == 0) { return; } @@ -143,7 +143,7 @@ inline void pack_list_data_impl( raft::device_matrix_view codes, std::variant offset_or_indices, uint32_t pq_bits, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { auto n_rows = codes.extent(0); if (n_rows == 0) { return; } diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh index d2bbbcf345..27690d64af 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh @@ -81,7 +81,7 @@ class InitializerGPU { size_t D; // Dimension size_t K; // Num of Centroids raft::resources const& handle_; // reusable resource handle - rmm::cuda_stream_view stream_ = + cuda::stream_ref stream_ = raft::resource::get_cuda_stream(handle_); // CUDA stream obtained from handle_ }; diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu index ca047b3246..03d8cc50ea 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu @@ -1037,7 +1037,7 @@ void IVFGPU::PrepareClusterSearchInputs( raft::device_vector& d_G_kbxSumq) { raft::resources const& searcher_handle = searcher.get_handle(); - rmm::cuda_stream_view searcher_stream = searcher.get_stream(); + cuda::stream_ref searcher_stream = searcher.get_stream(); const size_t batch_size = queries.extent(0); // Compute ||q - c||^2 = -2 * q . c + ||q||^2 + ||c||^2 into centroid_distances: diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh index 75ddaee865..4f8c94380c 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh @@ -342,7 +342,7 @@ class IVFGPU { void AllocateHostMemory(); raft::resources const& handle_; // reusable resource handle - rmm::cuda_stream_view stream_ = + cuda::stream_ref stream_ = raft::resource::get_cuda_stream(handle_); // CUDA stream obtained from handle_ // Device pointers for each data array. diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh index 90ef71ed12..55a78cd2f3 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh @@ -155,7 +155,7 @@ class DataQuantizerGPU { // RAFT resources raft::resources const& handle_; // reusable resource handle - rmm::cuda_stream_view stream_ = + cuda::stream_ref stream_ = raft::resource::get_cuda_stream(handle_); // CUDA stream obtained from handle_ // Device temporary buffers for quantization diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu index af5dfddb61..0dcc3826c2 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu @@ -115,7 +115,7 @@ void launchPrecomputeLUTs(const float* d_query, float* d_lut_for_queries, size_t num_queries, size_t D, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { // Launch precompute kernel dim3 gridDim(num_queries, 1, 1); diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh index 73557f57ea..1535af97b1 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh @@ -52,7 +52,7 @@ class SearcherGPU { // Getter methods std::string const& get_mode() { return mode_; } raft::resources const& get_handle() const { return handle_; } - rmm::cuda_stream_view get_stream() const { return stream_; } + cuda::stream_ref get_stream() const { return stream_; } float* get_centroid_distances() { return centroid_distances_.data_handle(); } float* get_q_norms() { return q_norms_.data_handle(); } @@ -110,7 +110,7 @@ class SearcherGPU { private: raft::resources const& handle_; // reusable resource handle - rmm::cuda_stream_view stream_ = + cuda::stream_ref stream_ = raft::resource::get_cuda_stream(handle_); // CUDA stream obtained from handle_ size_t D; // number of dimension const float* query_ = nullptr; // rotated query (non-owning) diff --git a/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh b/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh index 1fc32205df..c16117200e 100644 --- a/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh +++ b/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh @@ -149,7 +149,7 @@ void launch_kernel(const index& idx, float* out_distances, uint32_t* out_indices, uint32_t& grid_dim_x, - rmm::cuda_stream_view stream, + cuda::stream_ref stream, IvfSampleFilterT sample_filter) { static_assert(std::is_same_v, "IVF-SQ JIT-LTO scan only supports CodeT=uint8_t"); @@ -291,7 +291,7 @@ void ivf_sq_scan(raft::resources const& handle, uint32_t* out_indices, IvfSampleFilterT sample_filter, uint32_t& grid_dim_x, - rmm::cuda_stream_view stream) + cuda::stream_ref stream) { // Determine the fused top-k capacity (0 = disabled / fallback to materialization) int capacity = is_local_topk_feasible(k) ? raft::bound_by_power_of_two(int(k)) : 0; diff --git a/cpp/tests/neighbors/all_neighbors.cuh b/cpp/tests/neighbors/all_neighbors.cuh index 0b43023eff..f7a4018760 100644 --- a/cpp/tests/neighbors/all_neighbors.cuh +++ b/cpp/tests/neighbors/all_neighbors.cuh @@ -259,7 +259,7 @@ class AllNeighborsTest : public ::testing::TestWithParam { private: raft::device_resources_snmg handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AllNeighborsInputs ps; rmm::device_uvector database; }; diff --git a/cpp/tests/neighbors/ann_brute_force.cuh b/cpp/tests/neighbors/ann_brute_force.cuh index 3da3759804..363e2cfacc 100644 --- a/cpp/tests/neighbors/ann_brute_force.cuh +++ b/cpp/tests/neighbors/ann_brute_force.cuh @@ -158,7 +158,7 @@ class AnnBruteForceTest : public ::testing::TestWithParam ps; rmm::device_uvector database; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/ann_cagra.cuh b/cpp/tests/neighbors/ann_cagra.cuh index 96ad03c4e9..6c22cde826 100644 --- a/cpp/tests/neighbors/ann_cagra.cuh +++ b/cpp/tests/neighbors/ann_cagra.cuh @@ -567,7 +567,7 @@ class AnnCagraTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnCagraInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; @@ -772,7 +772,7 @@ class AnnCagraAddNodesTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnCagraInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; @@ -1153,7 +1153,7 @@ class AnnCagraFilterTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnCagraInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; @@ -1402,7 +1402,7 @@ class AnnCagraIndexFilteredMergeTest : public ::testing::TestWithParam database; rmm::device_uvector search_queries; @@ -1650,7 +1650,7 @@ class AnnCagraIndexMergeTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnCagraInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; @@ -2420,7 +2420,7 @@ class AnnCagraMultiPartitionTest : public ::testing::TestWithParam database; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/ann_hnsw_ace.cuh b/cpp/tests/neighbors/ann_hnsw_ace.cuh index 03e9746b9b..e0167502b9 100644 --- a/cpp/tests/neighbors/ann_hnsw_ace.cuh +++ b/cpp/tests/neighbors/ann_hnsw_ace.cuh @@ -660,7 +660,7 @@ class AnnHnswAceTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnHnswAceInputs ps; rmm::device_uvector database_dev; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/ann_ivf_flat.cuh b/cpp/tests/neighbors/ann_ivf_flat.cuh index a6f72e6d84..c4b22fff04 100644 --- a/cpp/tests/neighbors/ann_ivf_flat.cuh +++ b/cpp/tests/neighbors/ann_ivf_flat.cuh @@ -515,7 +515,7 @@ class AnnIVFFlatTest : public ::testing::TestWithParam> { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnIvfFlatInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/ann_ivf_pq.cuh b/cpp/tests/neighbors/ann_ivf_pq.cuh index 20e5231b71..3d002158d3 100644 --- a/cpp/tests/neighbors/ann_ivf_pq.cuh +++ b/cpp/tests/neighbors/ann_ivf_pq.cuh @@ -705,7 +705,7 @@ class ivf_pq_test : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; ivf_pq_inputs ps; // NOLINT rmm::device_uvector database; // NOLINT rmm::device_uvector search_queries; // NOLINT @@ -867,7 +867,7 @@ class ivf_pq_filter_test : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; ivf_pq_inputs ps; // NOLINT rmm::device_uvector database; // NOLINT rmm::device_uvector search_queries; // NOLINT diff --git a/cpp/tests/neighbors/ann_ivf_rabitq.cuh b/cpp/tests/neighbors/ann_ivf_rabitq.cuh index 938f41f846..9e9b4a13b4 100644 --- a/cpp/tests/neighbors/ann_ivf_rabitq.cuh +++ b/cpp/tests/neighbors/ann_ivf_rabitq.cuh @@ -247,7 +247,7 @@ class ivf_rabitq_test : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; ivf_rabitq_inputs ps; // NOLINT rmm::device_uvector database; // NOLINT rmm::device_uvector search_queries; // NOLINT diff --git a/cpp/tests/neighbors/ann_ivf_sq.cuh b/cpp/tests/neighbors/ann_ivf_sq.cuh index 2bd800493a..4d0d4af924 100644 --- a/cpp/tests/neighbors/ann_ivf_sq.cuh +++ b/cpp/tests/neighbors/ann_ivf_sq.cuh @@ -373,7 +373,7 @@ class AnnIVFSQTest : public ::testing::TestWithParam> { } raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnIvfSqInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/ann_nn_descent.cuh b/cpp/tests/neighbors/ann_nn_descent.cuh index a4474a9382..ef33d5a87a 100644 --- a/cpp/tests/neighbors/ann_nn_descent.cuh +++ b/cpp/tests/neighbors/ann_nn_descent.cuh @@ -181,7 +181,7 @@ class AnnNNDescentTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnNNDescentInputs ps; rmm::device_uvector database; }; @@ -339,7 +339,7 @@ class AnnNNDescentDistEpiTest : public ::testing::TestWithParam database; }; @@ -457,7 +457,7 @@ class AnnNNDescentBatchTest : public ::testing::TestWithParam database; }; diff --git a/cpp/tests/neighbors/ann_scann.cuh b/cpp/tests/neighbors/ann_scann.cuh index 81ef21c8e2..1f1095cca6 100644 --- a/cpp/tests/neighbors/ann_scann.cuh +++ b/cpp/tests/neighbors/ann_scann.cuh @@ -282,7 +282,7 @@ class scann_test : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; scann_inputs ps; // NOLINT rmm::device_uvector database; // NOLINT }; diff --git a/cpp/tests/neighbors/ann_vamana.cuh b/cpp/tests/neighbors/ann_vamana.cuh index d1b9b34864..eef3976b54 100644 --- a/cpp/tests/neighbors/ann_vamana.cuh +++ b/cpp/tests/neighbors/ann_vamana.cuh @@ -299,7 +299,7 @@ class AnnVamanaTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnVamanaInputs ps; rmm::device_uvector database; rmm::device_uvector search_queries; diff --git a/cpp/tests/neighbors/distance_nn.cu b/cpp/tests/neighbors/distance_nn.cu index 7c90c515d1..bd831a11a2 100644 --- a/cpp/tests/neighbors/distance_nn.cu +++ b/cpp/tests/neighbors/distance_nn.cu @@ -164,7 +164,7 @@ class NNTest : public ::testing::TestWithParam> { private: raft::resources handle; - rmm::cuda_stream_view stream; + cuda::stream_ref stream; NNInputs params_; ComparisonSummary summary; IdxT m; diff --git a/cpp/tests/neighbors/hnsw.cu b/cpp/tests/neighbors/hnsw.cu index ac86c19123..66a37b527d 100644 --- a/cpp/tests/neighbors/hnsw.cu +++ b/cpp/tests/neighbors/hnsw.cu @@ -159,7 +159,7 @@ class AnnHNSWTest : public ::testing::TestWithParam { private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnHNSWInputs ps; rmm::device_uvector database; rmm::device_uvector queries; diff --git a/cpp/tests/neighbors/refine.cu b/cpp/tests/neighbors/refine.cu index bf8a2bd98e..8804e6a0e0 100644 --- a/cpp/tests/neighbors/refine.cu +++ b/cpp/tests/neighbors/refine.cu @@ -86,7 +86,7 @@ class RefineTest : public ::testing::TestWithParam> { public: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; RefineHelper data; }; diff --git a/cpp/tests/neighbors/refine_helper.cuh b/cpp/tests/neighbors/refine_helper.cuh index 2d82021d34..a9dd7f3c58 100644 --- a/cpp/tests/neighbors/refine_helper.cuh +++ b/cpp/tests/neighbors/refine_helper.cuh @@ -127,7 +127,7 @@ class RefineHelper { public: RefineInputs p; const raft::resources& handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; raft::device_matrix dataset; raft::device_matrix queries; diff --git a/cpp/tests/neighbors/tiered_index.cu b/cpp/tests/neighbors/tiered_index.cu index 4d1fd9c18d..11ea38d6ec 100644 --- a/cpp/tests/neighbors/tiered_index.cu +++ b/cpp/tests/neighbors/tiered_index.cu @@ -220,7 +220,7 @@ class ANNTieredIndexTest : public ::testing::TestWithParam private: raft::resources handle_; - rmm::cuda_stream_view stream_; + cuda::stream_ref stream_; AnnTieredIndexInputs ps; rmm::device_uvector database; rmm::device_uvector queries; diff --git a/fern/pages/cpp_api/cpp-api-common-types-copy-serialization-and-utility-apis.md b/fern/pages/cpp_api/cpp-api-common-types-copy-serialization-and-utility-apis.md index eeffd82ae6..7349cf13d7 100644 --- a/fern/pages/cpp_api/cpp-api-common-types-copy-serialization-and-utility-apis.md +++ b/fern/pages/cpp_api/cpp-api-common-types-copy-serialization-and-utility-apis.md @@ -14,7 +14,7 @@ Asynchronously copies elements between compatible memory locations. ```cpp template void copy(OutputIterator dst, InputIterator src, SizeType n, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); ``` **Parameters** @@ -24,7 +24,7 @@ void copy(OutputIterator dst, InputIterator src, SizeType n, | `dst` | `OutputIterator` | Destination pointer or iterator. | | `src` | `InputIterator` | Source pointer or iterator. | | `n` | `SizeType` | Number of elements to copy. | -| `stream` | `rmm::cuda_stream_view` | CUDA stream used for the copy. | +| `stream` | `cuda::stream_ref` | CUDA stream used for the copy. | **Returns** @@ -39,7 +39,7 @@ Copies a dense matrix between compatible matrix views. ```cpp template -void copy_matrix(OutputView dst, InputView src, rmm::cuda_stream_view stream); +void copy_matrix(OutputView dst, InputView src, cuda::stream_ref stream); ``` **Parameters** @@ -48,7 +48,7 @@ void copy_matrix(OutputView dst, InputView src, rmm::cuda_stream_view stream); | --- | --- | --- | | `dst` | `OutputView` | Destination matrix view. | | `src` | `InputView` | Source matrix view. | -| `stream` | `rmm::cuda_stream_view` | CUDA stream used for the copy. | +| `stream` | `cuda::stream_ref` | CUDA stream used for the copy. | **Returns** @@ -64,7 +64,7 @@ Convenience helper for copying host data to device memory. ```cpp template void update_device(DevicePointer dst, HostPointer src, SizeType n, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); ``` **Parameters** @@ -74,7 +74,7 @@ void update_device(DevicePointer dst, HostPointer src, SizeType n, | `dst` | `DevicePointer` | Destination device pointer. | | `src` | `HostPointer` | Source host pointer. | | `n` | `SizeType` | Number of elements to copy. | -| `stream` | `rmm::cuda_stream_view` | CUDA stream used for the copy. | +| `stream` | `cuda::stream_ref` | CUDA stream used for the copy. | **Returns** @@ -90,7 +90,7 @@ Convenience helper for copying device data to host memory. ```cpp template void update_host(HostPointer dst, DevicePointer src, SizeType n, - rmm::cuda_stream_view stream); + cuda::stream_ref stream); ``` **Parameters** @@ -100,7 +100,7 @@ void update_host(HostPointer dst, DevicePointer src, SizeType n, | `dst` | `HostPointer` | Destination host pointer. | | `src` | `DevicePointer` | Source device pointer. | | `n` | `SizeType` | Number of elements to copy. | -| `stream` | `rmm::cuda_stream_view` | CUDA stream used for the copy. | +| `stream` | `cuda::stream_ref` | CUDA stream used for the copy. | **Returns** diff --git a/fern/pages/cpp_api/cpp-api-common-types-execution-resources.md b/fern/pages/cpp_api/cpp-api-common-types-execution-resources.md index 90fa3eba43..dbc603a6a8 100644 --- a/fern/pages/cpp_api/cpp-api-common-types-execution-resources.md +++ b/fern/pages/cpp_api/cpp-api-common-types-execution-resources.md @@ -25,7 +25,7 @@ _Source header: `raft/core/resource/cuda_stream.hpp`_ Returns the CUDA stream associated with a resources object. ```cpp -rmm::cuda_stream_view get_cuda_stream(raft::resources const& res); +cuda::stream_ref get_cuda_stream(raft::resources const& res); ``` **Parameters** @@ -36,7 +36,7 @@ rmm::cuda_stream_view get_cuda_stream(raft::resources const& res); **Returns** -`rmm::cuda_stream_view` +`cuda::stream_ref` #### raft::resource::sync_stream @@ -47,7 +47,7 @@ Synchronizes the CUDA stream associated with a resources object. ```cpp void sync_stream(raft::resources const& res); -void sync_stream(raft::resources const& res, rmm::cuda_stream_view stream); +void sync_stream(raft::resources const& res, cuda::stream_ref stream); ``` **Parameters** @@ -55,7 +55,7 @@ void sync_stream(raft::resources const& res, rmm::cuda_stream_view stream); | Name | Type | Description | | --- | --- | --- | | `res` | `raft::resources const&` | Resources object to synchronize. | -| `stream` | `rmm::cuda_stream_view` | Optional stream to synchronize instead of the main stream. | +| `stream` | `cuda::stream_ref` | Optional stream to synchronize instead of the main stream. | **Returns** @@ -92,7 +92,7 @@ _Source header: `raft/core/resource/cuda_stream_pool.hpp`_ Returns a stream from the configured stream pool. ```cpp -rmm::cuda_stream_view get_stream_from_stream_pool(raft::resources const& res); +cuda::stream_ref get_stream_from_stream_pool(raft::resources const& res); ``` **Parameters** @@ -103,7 +103,7 @@ rmm::cuda_stream_view get_stream_from_stream_pool(raft::resources const& res); **Returns** -`rmm::cuda_stream_view` +`cuda::stream_ref` #### raft::resource::sync_stream_pool @@ -210,7 +210,7 @@ Constructs a single-GPU resources object. ```cpp device_resources( - rmm::cuda_stream_view stream_view = cuda::stream_ref{cudaStreamPerThread}, + cuda::stream_ref stream_view = cuda::stream_ref{cudaStreamPerThread}, std::shared_ptr stream_pool = nullptr, std::shared_ptr workspace_resource = nullptr, std::optional allocation_limit = std::nullopt); @@ -220,7 +220,7 @@ device_resources( | Name | Type | Description | | --- | --- | --- | -| `stream_view` | `rmm::cuda_stream_view` | Default CUDA stream used by algorithms. | +| `stream_view` | `cuda::stream_ref` | Default CUDA stream used by algorithms. | | `stream_pool` | `std::shared_ptr` | Optional CUDA stream pool. | | `workspace_resource` | `std::shared_ptr` | Optional workspace memory resource. | | `allocation_limit` | `std::optional` | Optional temporary workspace allocation limit in bytes. | @@ -232,14 +232,14 @@ Synchronizes either the main stream or a specific CUDA stream. ```cpp void sync_stream() const; -void sync_stream(rmm::cuda_stream_view stream) const; +void sync_stream(cuda::stream_ref stream) const; ``` **Parameters** | Name | Type | Description | | --- | --- | --- | -| `stream` | `rmm::cuda_stream_view` | Stream to synchronize. Omit to synchronize the main stream. | +| `stream` | `cuda::stream_ref` | Stream to synchronize. Omit to synchronize the main stream. | **Returns** @@ -251,12 +251,12 @@ void sync_stream(rmm::cuda_stream_view stream) const; Returns the main CUDA stream associated with the resources object. ```cpp -rmm::cuda_stream_view get_stream() const; +cuda::stream_ref get_stream() const; ``` **Returns** -`rmm::cuda_stream_view` +`cuda::stream_ref` #### raft::device_resources::is_stream_pool_initialized @@ -290,8 +290,8 @@ rmm::cuda_stream_pool const& get_stream_pool() const; Returns a stream from the configured CUDA stream pool. ```cpp -rmm::cuda_stream_view get_stream_from_stream_pool() const; -rmm::cuda_stream_view get_stream_from_stream_pool(std::size_t stream_idx) const; +cuda::stream_ref get_stream_from_stream_pool() const; +cuda::stream_ref get_stream_from_stream_pool(std::size_t stream_idx) const; ``` **Parameters** @@ -302,7 +302,7 @@ rmm::cuda_stream_view get_stream_from_stream_pool(std::size_t stream_idx) const; **Returns** -`rmm::cuda_stream_view` +`cuda::stream_ref` #### raft::device_resources::get_next_usable_stream @@ -310,8 +310,8 @@ rmm::cuda_stream_view get_stream_from_stream_pool(std::size_t stream_idx) const; Returns a stream from the pool when one exists; otherwise returns the main stream. ```cpp -rmm::cuda_stream_view get_next_usable_stream() const; -rmm::cuda_stream_view get_next_usable_stream(std::size_t stream_idx) const; +cuda::stream_ref get_next_usable_stream() const; +cuda::stream_ref get_next_usable_stream(std::size_t stream_idx) const; ``` **Parameters** @@ -322,7 +322,7 @@ rmm::cuda_stream_view get_next_usable_stream(std::size_t stream_idx) const; **Returns** -`rmm::cuda_stream_view` +`cuda::stream_ref` #### raft::device_resources::sync_stream_pool diff --git a/fern/pages/other/resources.md b/fern/pages/other/resources.md index 9cc45196b5..883d3e1265 100644 --- a/fern/pages/other/resources.md +++ b/fern/pages/other/resources.md @@ -369,7 +369,7 @@ cudaStreamDestroy(stream); cudaStream_t stream; cudaStreamCreate(&stream); -raft::device_resources resources{rmm::cuda_stream_view{stream}}; +raft::device_resources resources{cuda::stream_ref{stream}}; // cuVS C++ calls using resources are enqueued on stream. diff --git a/fern/scripts/generate_api_reference.py b/fern/scripts/generate_api_reference.py index 9a02eb1d0d..93a0b2b9dd 100755 --- a/fern/scripts/generate_api_reference.py +++ b/fern/scripts/generate_api_reference.py @@ -820,9 +820,9 @@ def add_symbol( "raft-resource-get-cuda-stream", "raft::resource::get_cuda_stream", "Returns the CUDA stream associated with a resources object.", - "rmm::cuda_stream_view get_cuda_stream(raft::resources const& res);", + "cuda::stream_ref get_cuda_stream(raft::resources const& res);", [("res", "raft::resources const&", "Resources object to query.")], - "rmm::cuda_stream_view", + "cuda::stream_ref", nested=True, ) add_symbol( @@ -832,7 +832,7 @@ def add_symbol( "Synchronizes the CUDA stream associated with a resources object.", ( "void sync_stream(raft::resources const& res);\n" - "void sync_stream(raft::resources const& res, rmm::cuda_stream_view stream);" + "void sync_stream(raft::resources const& res, cuda::stream_ref stream);" ), [ ( @@ -842,7 +842,7 @@ def add_symbol( ), ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "Optional stream to synchronize instead of the main stream.", ), ], @@ -881,9 +881,9 @@ def add_symbol( "raft-resource-get-stream-from-stream-pool", "raft::resource::get_stream_from_stream_pool", "Returns a stream from the configured stream pool.", - "rmm::cuda_stream_view get_stream_from_stream_pool(raft::resources const& res);", + "cuda::stream_ref get_stream_from_stream_pool(raft::resources const& res);", [("res", "raft::resources const&", "Resources object to query.")], - "rmm::cuda_stream_view", + "cuda::stream_ref", nested=True, ) add_symbol( @@ -967,7 +967,7 @@ def add_symbol( "Constructs a single-GPU resources object.", ( "device_resources(\n" - " rmm::cuda_stream_view stream_view = cuda::stream_ref{cudaStreamPerThread},\n" + " cuda::stream_ref stream_view = cuda::stream_ref{cudaStreamPerThread},\n" " std::shared_ptr stream_pool = nullptr,\n" " std::shared_ptr workspace_resource = nullptr,\n" " std::optional allocation_limit = std::nullopt);" @@ -975,7 +975,7 @@ def add_symbol( [ ( "stream_view", - "rmm::cuda_stream_view", + "cuda::stream_ref", "Default CUDA stream used by algorithms.", ), ( @@ -1002,12 +1002,12 @@ def add_symbol( "Synchronizes either the main stream or a specific CUDA stream.", ( "void sync_stream() const;\n" - "void sync_stream(rmm::cuda_stream_view stream) const;" + "void sync_stream(cuda::stream_ref stream) const;" ), [ ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "Stream to synchronize. Omit to synchronize the main stream.", ) ], @@ -1018,8 +1018,8 @@ def add_symbol( "raft-device-resources-get-stream", "raft::device_resources::get_stream", "Returns the main CUDA stream associated with the resources object.", - "rmm::cuda_stream_view get_stream() const;", - returns="rmm::cuda_stream_view", + "cuda::stream_ref get_stream() const;", + returns="cuda::stream_ref", ) add_symbol( lines, @@ -1043,8 +1043,8 @@ def add_symbol( "raft::device_resources::get_stream_from_stream_pool", "Returns a stream from the configured CUDA stream pool.", ( - "rmm::cuda_stream_view get_stream_from_stream_pool() const;\n" - "rmm::cuda_stream_view get_stream_from_stream_pool(std::size_t stream_idx) const;" + "cuda::stream_ref get_stream_from_stream_pool() const;\n" + "cuda::stream_ref get_stream_from_stream_pool(std::size_t stream_idx) const;" ), [ ( @@ -1053,7 +1053,7 @@ def add_symbol( "Optional index of the stream in the stream pool.", ) ], - "rmm::cuda_stream_view", + "cuda::stream_ref", ) add_symbol( lines, @@ -1064,8 +1064,8 @@ def add_symbol( "the main stream." ), ( - "rmm::cuda_stream_view get_next_usable_stream() const;\n" - "rmm::cuda_stream_view get_next_usable_stream(std::size_t stream_idx) const;" + "cuda::stream_ref get_next_usable_stream() const;\n" + "cuda::stream_ref get_next_usable_stream(std::size_t stream_idx) const;" ), [ ( @@ -1074,7 +1074,7 @@ def add_symbol( "Optional stream pool index to use when a stream pool is configured.", ) ], - "rmm::cuda_stream_view", + "cuda::stream_ref", ) add_symbol( lines, @@ -2105,7 +2105,7 @@ def add_symbol( ( "template \n" "void copy(OutputIterator dst, InputIterator src, SizeType n,\n" - " rmm::cuda_stream_view stream);" + " cuda::stream_ref stream);" ), [ ("dst", "OutputIterator", "Destination pointer or iterator."), @@ -2113,7 +2113,7 @@ def add_symbol( ("n", "SizeType", "Number of elements to copy."), ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "CUDA stream used for the copy.", ), ], @@ -2124,13 +2124,13 @@ def add_symbol( "raft-copy-matrix", "raft::copy_matrix", "Copies a dense matrix between compatible matrix views.", - "template \nvoid copy_matrix(OutputView dst, InputView src, rmm::cuda_stream_view stream);", + "template \nvoid copy_matrix(OutputView dst, InputView src, cuda::stream_ref stream);", [ ("dst", "OutputView", "Destination matrix view."), ("src", "InputView", "Source matrix view."), ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "CUDA stream used for the copy.", ), ], @@ -2144,7 +2144,7 @@ def add_symbol( ( "template \n" "void update_device(DevicePointer dst, HostPointer src, SizeType n,\n" - " rmm::cuda_stream_view stream);" + " cuda::stream_ref stream);" ), [ ("dst", "DevicePointer", "Destination device pointer."), @@ -2152,7 +2152,7 @@ def add_symbol( ("n", "SizeType", "Number of elements to copy."), ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "CUDA stream used for the copy.", ), ], @@ -2166,7 +2166,7 @@ def add_symbol( ( "template \n" "void update_host(HostPointer dst, DevicePointer src, SizeType n,\n" - " rmm::cuda_stream_view stream);" + " cuda::stream_ref stream);" ), [ ("dst", "HostPointer", "Destination host pointer."), @@ -2174,7 +2174,7 @@ def add_symbol( ("n", "SizeType", "Number of elements to copy."), ( "stream", - "rmm::cuda_stream_view", + "cuda::stream_ref", "CUDA stream used for the copy.", ), ], From b452761a35fb442159ec5ec6ac513f282152e91f Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Wed, 2 Sep 2026 17:34:32 -0500 Subject: [PATCH 2/5] Use cuda::stream_ref --- c/src/core/c_api.cpp | 2 +- c/tests/neighbors/ann_mg_c.cu | 1 - cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h | 4 ++-- cpp/include/cuvs/neighbors/scann.hpp | 3 +-- cpp/include/cuvs/neighbors/vamana.hpp | 1 - cpp/internal/cuvs_internal/neighbors/refine_helper.cuh | 4 ++-- cpp/src/neighbors/detail/ann_utils.cuh | 2 +- cpp/src/neighbors/detail/cagra/compute_distance.hpp | 1 + .../neighbors/detail/cagra/compute_distance_standard-impl.cuh | 1 + cpp/src/neighbors/detail/cagra/compute_distance_vpq-impl.cuh | 1 + cpp/src/neighbors/detail/cagra/search_plan.cuh | 1 + cpp/src/neighbors/ivf_common.cu | 1 + cpp/src/neighbors/ivf_common.cuh | 1 + .../ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh | 3 ++- cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh | 3 ++- cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh | 2 +- .../ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in | 3 ++- .../neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu | 3 ++- cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu | 3 ++- cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh | 2 +- cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh | 4 ++-- cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh | 2 +- cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh | 3 ++- cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh | 1 + cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp | 3 ++- cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh | 1 + cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh | 1 + cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu | 1 + cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh | 4 ++-- cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh | 3 ++- cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu | 1 + cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh | 4 ++-- cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh | 1 + cpp/tests/neighbors/all_neighbors.cuh | 1 + cpp/tests/neighbors/ann_brute_force.cuh | 1 + cpp/tests/neighbors/ann_cagra.cuh | 1 + cpp/tests/neighbors/ann_hnsw_ace.cuh | 1 + cpp/tests/neighbors/ann_ivf_flat.cuh | 1 + cpp/tests/neighbors/ann_ivf_pq.cuh | 1 + cpp/tests/neighbors/ann_ivf_rabitq.cuh | 3 ++- cpp/tests/neighbors/ann_ivf_sq.cuh | 1 + cpp/tests/neighbors/ann_nn_descent.cuh | 1 + cpp/tests/neighbors/ann_scann.cuh | 1 + cpp/tests/neighbors/ann_vamana.cuh | 1 + cpp/tests/neighbors/distance_nn.cu | 1 + cpp/tests/neighbors/hnsw.cu | 1 + cpp/tests/neighbors/refine.cu | 4 ++-- cpp/tests/neighbors/refine_helper.cuh | 4 ++-- cpp/tests/neighbors/tiered_index.cu | 1 + fern/pages/other/resources.md | 2 +- 50 files changed, 64 insertions(+), 33 deletions(-) diff --git a/c/src/core/c_api.cpp b/c/src/core/c_api.cpp index ca4e7fcd90..4b5c2c351c 100644 --- a/c/src/core/c_api.cpp +++ b/c/src/core/c_api.cpp @@ -15,7 +15,7 @@ #include #include #include -#include +#include #include #include #include diff --git a/c/tests/neighbors/ann_mg_c.cu b/c/tests/neighbors/ann_mg_c.cu index f9d651e387..7f8fb7da24 100644 --- a/c/tests/neighbors/ann_mg_c.cu +++ b/c/tests/neighbors/ann_mg_c.cu @@ -6,7 +6,6 @@ #include #include #include - #include #include #include diff --git a/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h b/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h index 64e43412dd..5600b58af2 100644 --- a/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h +++ b/cpp/bench/ann/src/cuvs/cuvs_ann_bench_utils.h @@ -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 */ #pragma once @@ -18,7 +18,7 @@ #include #include -#include +#include #include #include #include diff --git a/cpp/include/cuvs/neighbors/scann.hpp b/cpp/include/cuvs/neighbors/scann.hpp index de0487265e..f186ef21e2 100644 --- a/cpp/include/cuvs/neighbors/scann.hpp +++ b/cpp/include/cuvs/neighbors/scann.hpp @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -17,7 +17,6 @@ #include #include #include -#include #include #include diff --git a/cpp/include/cuvs/neighbors/vamana.hpp b/cpp/include/cuvs/neighbors/vamana.hpp index 398c8cd6e5..517e1f76e1 100644 --- a/cpp/include/cuvs/neighbors/vamana.hpp +++ b/cpp/include/cuvs/neighbors/vamana.hpp @@ -16,7 +16,6 @@ #include #include #include -#include #include #include diff --git a/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh b/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh index 3aa7b78bb9..f5cfe6b79d 100644 --- a/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh +++ b/cpp/internal/cuvs_internal/neighbors/refine_helper.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2023, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -15,7 +15,7 @@ #include #include -#include +#include #include namespace cuvs::neighbors { diff --git a/cpp/src/neighbors/detail/ann_utils.cuh b/cpp/src/neighbors/detail/ann_utils.cuh index a803ed3ca5..deddc47dcb 100644 --- a/cpp/src/neighbors/detail/ann_utils.cuh +++ b/cpp/src/neighbors/detail/ann_utils.cuh @@ -18,7 +18,7 @@ #include #include -#include +#include #include #include #include diff --git a/cpp/src/neighbors/detail/cagra/compute_distance.hpp b/cpp/src/neighbors/detail/cagra/compute_distance.hpp index f080932cdc..c1d0033a01 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance.hpp +++ b/cpp/src/neighbors/detail/cagra/compute_distance.hpp @@ -9,6 +9,7 @@ #include "hashmap.hpp" #include "utils.hpp" +#include #include #include #include diff --git a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh index 59b40167b3..0c8e8fdff9 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh +++ b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh @@ -6,6 +6,7 @@ #include "compute_distance_standard.hpp" +#include #include #include diff --git a/cpp/src/neighbors/detail/cagra/compute_distance_vpq-impl.cuh b/cpp/src/neighbors/detail/cagra/compute_distance_vpq-impl.cuh index e42fa6e2b7..3dbd17859b 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance_vpq-impl.cuh +++ b/cpp/src/neighbors/detail/cagra/compute_distance_vpq-impl.cuh @@ -8,6 +8,7 @@ #include "compute_distance_vpq.hpp" #include "packed_type.hpp" +#include #include #include diff --git a/cpp/src/neighbors/detail/cagra/search_plan.cuh b/cpp/src/neighbors/detail/cagra/search_plan.cuh index 8ce5d78cfe..da42f637dc 100644 --- a/cpp/src/neighbors/detail/cagra/search_plan.cuh +++ b/cpp/src/neighbors/detail/cagra/search_plan.cuh @@ -7,6 +7,7 @@ #include "hashmap.hpp" +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_common.cu b/cpp/src/neighbors/ivf_common.cu index ec7da858d8..a7789a2ffa 100644 --- a/cpp/src/neighbors/ivf_common.cu +++ b/cpp/src/neighbors/ivf_common.cu @@ -5,6 +5,7 @@ #include "ivf_common.cuh" +#include #include #include diff --git a/cpp/src/neighbors/ivf_common.cuh b/cpp/src/neighbors/ivf_common.cuh index 9dd3cd15bb..c445e3b6b7 100644 --- a/cpp/src/neighbors/ivf_common.cuh +++ b/cpp/src/neighbors/ivf_common.cuh @@ -5,6 +5,7 @@ #pragma once +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh index f5f61bc786..3dd41c6fda 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_explicit_inst.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -8,6 +8,7 @@ #include "../detail/ann_utils.cuh" #include "ivf_flat_interleaved_scan_jit.cuh" #include +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh index e5d8da76d6..199303d151 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_ext.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,6 +11,7 @@ #include #include "../detail/ann_utils.cuh" +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh index c9bfc93e7f..48aa62501c 100644 --- a/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh +++ b/cpp/src/neighbors/ivf_flat/ivf_flat_interleaved_scan_jit.cuh @@ -25,7 +25,7 @@ #include // RAFT_CUDA_TRY #include -#include +#include namespace cuvs::neighbors::ivf_flat::detail { static constexpr int kThreadsPerBlock = 128; diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in index 320f6bd98e..30de9d57de 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_compute_similarity_run_inst.cu.in @@ -1,8 +1,9 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ +#include #include #include diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu index d215c69202..6900dd5e53 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_contiguous_list_data.cu @@ -1,9 +1,10 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include "../ivf_pq_contiguous_list_data_impl.cuh" +#include #include namespace cuvs::neighbors::ivf_pq::detail { diff --git a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu index c439298f04..1fae9b28d7 100644 --- a/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu +++ b/cpp/src/neighbors/ivf_pq/detail/ivf_pq_list_data.cu @@ -1,9 +1,10 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #include "../ivf_pq_list_data_impl.cuh" +#include #include namespace cuvs::neighbors::ivf_pq::detail { diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh index 187b382b9f..aaa040e6d2 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_build.cuh @@ -50,7 +50,7 @@ #include #include -#include +#include #include #include diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh index 34e9b760eb..d313749840 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity.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 */ @@ -9,12 +9,12 @@ #include "ivf_pq_fp_8bit.cuh" // cuvs::neighbors::ivf_pq::detail::fp_8bit #include "ivf_pq_compute_similarity.hpp" // cuvs::neighbors::ivf_pq::detail::selected +#include // cuda::stream_ref #include #include // cuvs::distance::DistanceType #include #include // cuvs::neighbors::ivf_pq::codebook_gen #include // RAFT_WEAK_FUNCTION -#include // cuda::stream_ref #include // __half diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh index 212e596cae..37bfa550d4 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_compute_similarity_impl.cuh @@ -21,7 +21,7 @@ #include // raft::Pow2 #include // raft::TxN_t -#include // cuda::stream_ref +#include // cuda::stream_ref namespace cuvs::neighbors::ivf_pq::detail { diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh index 507528e905..59e5ff777c 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data.cuh @@ -1,10 +1,11 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once #include +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh index 369f9ad31c..0e347526b4 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_contiguous_list_data_impl.cuh @@ -6,6 +6,7 @@ #pragma once #include "ivf_pq_codepacking.cuh" #include +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp index 774a2a4ef5..7a06444732 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data.hpp @@ -1,10 +1,11 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once #include +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh index ab2df79966..88240756af 100644 --- a/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh +++ b/cpp/src/neighbors/ivf_pq/ivf_pq_list_data_impl.cuh @@ -5,6 +5,7 @@ #pragma once #include "ivf_pq_codepacking.cuh" +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh index 27690d64af..3d76778290 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/initializer_gpu.cuh @@ -13,6 +13,7 @@ #include +#include #include #include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu index 03d8cc50ea..1f24f3715c 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cu @@ -12,6 +12,7 @@ #include "../utils/reductions.cuh" #include "ivf_gpu.cuh" #include "searcher_gpu.cuh" +#include #include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh index 4f8c94380c..0c0c1695d7 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/ivf_gpu.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -20,7 +20,7 @@ #include #include -#include +#include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh index 55a78cd2f3..39e389dd9c 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/quantizer_gpu.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -11,6 +11,7 @@ #include "../defines.hpp" #include "rotator_gpu.cuh" +#include #include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu index 0dcc3826c2..1e8608da3f 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cu @@ -16,6 +16,7 @@ #include "searcher_gpu.cuh" #include "searcher_gpu_common.cuh" +#include #include #include diff --git a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh index 1535af97b1..3d5cd8c350 100644 --- a/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh +++ b/cpp/src/neighbors/ivf_rabitq/gpu_index/searcher_gpu.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -17,7 +17,7 @@ #include #include -#include +#include namespace cuvs::neighbors::ivf_rabitq::detail { diff --git a/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh b/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh index c16117200e..716e7b733c 100644 --- a/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh +++ b/cpp/src/neighbors/ivf_sq/ivf_sq_search.cuh @@ -16,6 +16,7 @@ #include #include +#include #include #include #include diff --git a/cpp/tests/neighbors/all_neighbors.cuh b/cpp/tests/neighbors/all_neighbors.cuh index f7a4018760..8a5b435982 100644 --- a/cpp/tests/neighbors/all_neighbors.cuh +++ b/cpp/tests/neighbors/all_neighbors.cuh @@ -10,6 +10,7 @@ #include "ann_utils.cuh" #include "naive_knn.cuh" #include +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_brute_force.cuh b/cpp/tests/neighbors/ann_brute_force.cuh index 363e2cfacc..c3feaaec69 100644 --- a/cpp/tests/neighbors/ann_brute_force.cuh +++ b/cpp/tests/neighbors/ann_brute_force.cuh @@ -11,6 +11,7 @@ #include +#include #include #include diff --git a/cpp/tests/neighbors/ann_cagra.cuh b/cpp/tests/neighbors/ann_cagra.cuh index 6c22cde826..381bddb989 100644 --- a/cpp/tests/neighbors/ann_cagra.cuh +++ b/cpp/tests/neighbors/ann_cagra.cuh @@ -7,6 +7,7 @@ #include "../test_utils.cuh" #include "ann_utils.cuh" #include "vpq_utils.cuh" +#include #include #include "cagra_padded_build_helpers.cuh" diff --git a/cpp/tests/neighbors/ann_hnsw_ace.cuh b/cpp/tests/neighbors/ann_hnsw_ace.cuh index e0167502b9..f953da0c55 100644 --- a/cpp/tests/neighbors/ann_hnsw_ace.cuh +++ b/cpp/tests/neighbors/ann_hnsw_ace.cuh @@ -9,6 +9,7 @@ #include #include +#include #include #include diff --git a/cpp/tests/neighbors/ann_ivf_flat.cuh b/cpp/tests/neighbors/ann_ivf_flat.cuh index c4b22fff04..47a721bf2f 100644 --- a/cpp/tests/neighbors/ann_ivf_flat.cuh +++ b/cpp/tests/neighbors/ann_ivf_flat.cuh @@ -8,6 +8,7 @@ #include "ann_utils.cuh" #include "naive_knn.cuh" +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_ivf_pq.cuh b/cpp/tests/neighbors/ann_ivf_pq.cuh index 3d002158d3..85ec62678f 100644 --- a/cpp/tests/neighbors/ann_ivf_pq.cuh +++ b/cpp/tests/neighbors/ann_ivf_pq.cuh @@ -11,6 +11,7 @@ #include #include +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_ivf_rabitq.cuh b/cpp/tests/neighbors/ann_ivf_rabitq.cuh index 9e9b4a13b4..cdf0165b15 100644 --- a/cpp/tests/neighbors/ann_ivf_rabitq.cuh +++ b/cpp/tests/neighbors/ann_ivf_rabitq.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2025-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -7,6 +7,7 @@ #include "../test_utils.cuh" #include "ann_utils.cuh" #include "naive_knn.cuh" +#include #include #include diff --git a/cpp/tests/neighbors/ann_ivf_sq.cuh b/cpp/tests/neighbors/ann_ivf_sq.cuh index 4d0d4af924..5299efd920 100644 --- a/cpp/tests/neighbors/ann_ivf_sq.cuh +++ b/cpp/tests/neighbors/ann_ivf_sq.cuh @@ -8,6 +8,7 @@ #include "ann_utils.cuh" #include "naive_knn.cuh" +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_nn_descent.cuh b/cpp/tests/neighbors/ann_nn_descent.cuh index ef33d5a87a..7fd03b2f43 100644 --- a/cpp/tests/neighbors/ann_nn_descent.cuh +++ b/cpp/tests/neighbors/ann_nn_descent.cuh @@ -14,6 +14,7 @@ #include #include +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_scann.cuh b/cpp/tests/neighbors/ann_scann.cuh index 1f1095cca6..8f6deb30de 100644 --- a/cpp/tests/neighbors/ann_scann.cuh +++ b/cpp/tests/neighbors/ann_scann.cuh @@ -12,6 +12,7 @@ #include #include +#include #include #include #include diff --git a/cpp/tests/neighbors/ann_vamana.cuh b/cpp/tests/neighbors/ann_vamana.cuh index eef3976b54..ada130634a 100644 --- a/cpp/tests/neighbors/ann_vamana.cuh +++ b/cpp/tests/neighbors/ann_vamana.cuh @@ -7,6 +7,7 @@ #include "../test_utils.cuh" #include "ann_utils.cuh" +#include #include #include "cagra_padded_build_helpers.cuh" diff --git a/cpp/tests/neighbors/distance_nn.cu b/cpp/tests/neighbors/distance_nn.cu index bd831a11a2..c7018b77c9 100644 --- a/cpp/tests/neighbors/distance_nn.cu +++ b/cpp/tests/neighbors/distance_nn.cu @@ -9,6 +9,7 @@ #include "../../src/distance/fused_distance_nn.cuh" #include "../../src/distance/unfused_distance_nn.cuh" +#include #include #include #include diff --git a/cpp/tests/neighbors/hnsw.cu b/cpp/tests/neighbors/hnsw.cu index 66a37b527d..01a30950d0 100644 --- a/cpp/tests/neighbors/hnsw.cu +++ b/cpp/tests/neighbors/hnsw.cu @@ -8,6 +8,7 @@ #include "cagra_padded_build_helpers.cuh" #include +#include #include #include #include diff --git a/cpp/tests/neighbors/refine.cu b/cpp/tests/neighbors/refine.cu index 8804e6a0e0..1033fd6221 100644 --- a/cpp/tests/neighbors/refine.cu +++ b/cpp/tests/neighbors/refine.cu @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2025, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ @@ -14,7 +14,7 @@ #include #include -#include +#include #include diff --git a/cpp/tests/neighbors/refine_helper.cuh b/cpp/tests/neighbors/refine_helper.cuh index a9dd7f3c58..1680d9b5cf 100644 --- a/cpp/tests/neighbors/refine_helper.cuh +++ b/cpp/tests/neighbors/refine_helper.cuh @@ -1,5 +1,5 @@ /* - * SPDX-FileCopyrightText: Copyright (c) 2022-2024, NVIDIA CORPORATION. + * SPDX-FileCopyrightText: Copyright (c) 2022-2026, NVIDIA CORPORATION & AFFILIATES. All rights reserved. * SPDX-License-Identifier: Apache-2.0 */ #pragma once @@ -15,7 +15,7 @@ #include "naive_knn.cuh" -#include +#include #include namespace cuvs::neighbors { diff --git a/cpp/tests/neighbors/tiered_index.cu b/cpp/tests/neighbors/tiered_index.cu index 11ea38d6ec..5c71c31319 100644 --- a/cpp/tests/neighbors/tiered_index.cu +++ b/cpp/tests/neighbors/tiered_index.cu @@ -7,6 +7,7 @@ #include "ann_utils.cuh" #include +#include #include #include #include diff --git a/fern/pages/other/resources.md b/fern/pages/other/resources.md index 883d3e1265..5b3c808e95 100644 --- a/fern/pages/other/resources.md +++ b/fern/pages/other/resources.md @@ -362,7 +362,7 @@ cudaStreamDestroy(stream); ```cpp #include -#include +#include #include From 52fe5f979a1c73afd32484315abe1f574b4f2bb0 Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Sat, 5 Sep 2026 21:43:14 -0500 Subject: [PATCH 3/5] Remove duplicate list size initialization --- cpp/src/neighbors/ivf_sq_index.cpp | 2 -- 1 file changed, 2 deletions(-) diff --git a/cpp/src/neighbors/ivf_sq_index.cpp b/cpp/src/neighbors/ivf_sq_index.cpp index b66ef0d709..14733f75d8 100644 --- a/cpp/src/neighbors/ivf_sq_index.cpp +++ b/cpp/src/neighbors/ivf_sq_index.cpp @@ -46,8 +46,6 @@ index::index(raft::resources const& res, check_consistency(); auto stream = raft::resource::get_cuda_stream(res); std::memset(accum_sorted_sizes_.data_handle(), 0, accum_sorted_sizes_.size() * sizeof(int64_t)); - RAFT_CUDA_TRY(cudaMemsetAsync( - list_sizes_.data_handle(), 0, list_sizes_.size() * sizeof(uint32_t), stream.get())); RAFT_CUDA_TRY( cudaMemsetAsync(data_ptrs_.data_handle(), 0, data_ptrs_.size() * sizeof(CodeT*), stream.get())); RAFT_CUDA_TRY(cudaMemsetAsync( From 493302246b20de44b2f84146701b6f5fa93dff36 Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Sun, 6 Sep 2026 09:32:46 -0500 Subject: [PATCH 4/5] Apply migration style checks --- .../cagra/compute_distance_standard-impl.cuh | 34 +++++++++---------- 1 file changed, 17 insertions(+), 17 deletions(-) diff --git a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh index 0c8e8fdff9..b9061f8796 100644 --- a/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh +++ b/cpp/src/neighbors/detail/cagra/compute_distance_standard-impl.cuh @@ -132,23 +132,23 @@ standard_descriptor_spec* dev_ptr, - cuda::stream_ref stream) { - standard_dataset_descriptor_init_kernel - <<<1, 1, 0, stream.get()>>>(dev_ptr, ptr, size, dim, ld, dataset_norms); - RAFT_CUDA_TRY(cudaPeekAtLastError()); - }, - Metric, - DatasetBlockDim, - false, // is_vpq - 0, // pq_bits - 0}; // pq_len + return host_type{ + desc_type{ptr, size, dim, ld, dataset_norms}, + [=](dataset_descriptor_base_t* dev_ptr, cuda::stream_ref stream) { + standard_dataset_descriptor_init_kernel + <<<1, 1, 0, stream.get()>>>(dev_ptr, ptr, size, dim, ld, dataset_norms); + RAFT_CUDA_TRY(cudaPeekAtLastError()); + }, + Metric, + DatasetBlockDim, + false, // is_vpq + 0, // pq_bits + 0}; // pq_len } } // namespace cuvs::neighbors::cagra::detail From a2eea53805152209ade7fcc411d74a1877b679f7 Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Sun, 6 Sep 2026 15:17:42 -0500 Subject: [PATCH 5/5] Restore IVF-SQ list size initialization --- c/tests/neighbors/ann_mg_c.cu | 1 + cpp/src/neighbors/ivf_sq_index.cpp | 2 ++ 2 files changed, 3 insertions(+) diff --git a/c/tests/neighbors/ann_mg_c.cu b/c/tests/neighbors/ann_mg_c.cu index 7f8fb7da24..f9d651e387 100644 --- a/c/tests/neighbors/ann_mg_c.cu +++ b/c/tests/neighbors/ann_mg_c.cu @@ -6,6 +6,7 @@ #include #include #include + #include #include #include diff --git a/cpp/src/neighbors/ivf_sq_index.cpp b/cpp/src/neighbors/ivf_sq_index.cpp index 14733f75d8..b66ef0d709 100644 --- a/cpp/src/neighbors/ivf_sq_index.cpp +++ b/cpp/src/neighbors/ivf_sq_index.cpp @@ -46,6 +46,8 @@ index::index(raft::resources const& res, check_consistency(); auto stream = raft::resource::get_cuda_stream(res); std::memset(accum_sorted_sizes_.data_handle(), 0, accum_sorted_sizes_.size() * sizeof(int64_t)); + RAFT_CUDA_TRY(cudaMemsetAsync( + list_sizes_.data_handle(), 0, list_sizes_.size() * sizeof(uint32_t), stream.get())); RAFT_CUDA_TRY( cudaMemsetAsync(data_ptrs_.data_handle(), 0, data_ptrs_.size() * sizeof(CodeT*), stream.get())); RAFT_CUDA_TRY(cudaMemsetAsync(