From ac53a0fdc35fe36a11a3ac4debd4cc9a4076fe7f Mon Sep 17 00:00:00 2001 From: Micka Date: Tue, 10 Sep 2024 17:20:52 +0200 Subject: [PATCH 1/3] [BUG] Fix bitset function visibility (#2429) `raft::ceildiv` is also being replaced with `raft::div_rounding_up_safe` to avoid including CUDA headers when not needed. Authors: - Micka (https://github.com/lowener) Approvers: - rhdong (https://github.com/rhdong) - Corey J. Nolet (https://github.com/cjnolet) URL: https://github.com/rapidsai/raft/pull/2429 --- cpp/include/raft/core/bitmap.cuh | 6 +++--- cpp/include/raft/core/bitset.cuh | 24 ++++++------------------ cpp/include/raft/core/bitset.hpp | 11 +++++++++-- 3 files changed, 18 insertions(+), 23 deletions(-) diff --git a/cpp/include/raft/core/bitmap.cuh b/cpp/include/raft/core/bitmap.cuh index cafd1977ab..024b1244a6 100644 --- a/cpp/include/raft/core/bitmap.cuh +++ b/cpp/include/raft/core/bitmap.cuh @@ -35,9 +35,9 @@ _RAFT_HOST_DEVICE inline bool bitmap_view::test(const index_t } template -_RAFT_HOST_DEVICE void bitmap_view::set(const index_t row, - const index_t col, - bool new_value) const +_RAFT_DEVICE void bitmap_view::set(const index_t row, + const index_t col, + bool new_value) const { set(row * cols_ + col, new_value); } diff --git a/cpp/include/raft/core/bitset.cuh b/cpp/include/raft/core/bitset.cuh index 0cdb4c1fb6..b6e6128eca 100644 --- a/cpp/include/raft/core/bitset.cuh +++ b/cpp/include/raft/core/bitset.cuh @@ -46,8 +46,8 @@ _RAFT_HOST_DEVICE bool bitset_view::operator[](const index_t } template -_RAFT_HOST_DEVICE void bitset_view::set(const index_t sample_index, - bool set_value) const +_RAFT_DEVICE void bitset_view::set(const index_t sample_index, + bool set_value) const { const index_t bit_element = sample_index / bitset_element_size; const index_t bit_index = sample_index % bitset_element_size; @@ -60,18 +60,12 @@ _RAFT_HOST_DEVICE void bitset_view::set(const index_t sample_ } } -template -_RAFT_HOST_DEVICE inline index_t bitset_view::n_elements() const -{ - return raft::ceildiv(bitset_len_, bitset_element_size); -} - template bitset::bitset(const raft::resources& res, raft::device_vector_view mask_index, index_t bitset_len, bool default_value) - : bitset_{std::size_t(raft::ceildiv(bitset_len, bitset_element_size)), + : bitset_{std::size_t(raft::div_rounding_up_safe(bitset_len, bitset_element_size)), raft::resource::get_cuda_stream(res)}, bitset_len_{bitset_len} { @@ -83,26 +77,20 @@ template bitset::bitset(const raft::resources& res, index_t bitset_len, bool default_value) - : bitset_{std::size_t(raft::ceildiv(bitset_len, bitset_element_size)), + : bitset_{std::size_t(raft::div_rounding_up_safe(bitset_len, bitset_element_size)), raft::resource::get_cuda_stream(res)}, bitset_len_{bitset_len} { reset(res, default_value); } -template -index_t bitset::n_elements() const -{ - return raft::ceildiv(bitset_len_, bitset_element_size); -} - template void bitset::resize(const raft::resources& res, index_t new_bitset_len, bool default_value) { - auto old_size = raft::ceildiv(bitset_len_, bitset_element_size); - auto new_size = raft::ceildiv(new_bitset_len, bitset_element_size); + auto old_size = raft::div_rounding_up_safe(bitset_len_, bitset_element_size); + auto new_size = raft::div_rounding_up_safe(new_bitset_len, bitset_element_size); bitset_.resize(new_size); bitset_len_ = new_bitset_len; if (old_size < new_size) { diff --git a/cpp/include/raft/core/bitset.hpp b/cpp/include/raft/core/bitset.hpp index 0df12f25e6..3608ee43fa 100644 --- a/cpp/include/raft/core/bitset.hpp +++ b/cpp/include/raft/core/bitset.hpp @@ -20,6 +20,7 @@ #include #include #include +#include namespace raft::core { /** @@ -89,7 +90,10 @@ struct bitset_view { /** * @brief Get the number of elements used by the bitset representation. */ - inline _RAFT_HOST_DEVICE auto n_elements() const -> index_t; + inline _RAFT_HOST_DEVICE auto n_elements() const -> index_t + { + return raft::div_rounding_up_safe(bitset_len_, bitset_element_size); + } inline auto to_mdspan() -> raft::device_vector_view { @@ -173,7 +177,10 @@ struct bitset { /** * @brief Get the number of elements used by the bitset representation. */ - inline auto n_elements() const -> index_t; + inline auto n_elements() const -> index_t + { + return raft::div_rounding_up_safe(bitset_len_, bitset_element_size); + } /** @brief Get an mdspan view of the current bitset */ inline auto to_mdspan() -> raft::device_vector_view From 20718412dfa2026b94de2de2e822b73fe26a3bbc Mon Sep 17 00:00:00 2001 From: Ben Frederickson Date: Tue, 17 Sep 2024 11:56:06 -0700 Subject: [PATCH 2/3] Allow coo_sort to work on int64_t indices (#2432) Authors: - Ben Frederickson (https://github.com/benfred) Approvers: - Divye Gala (https://github.com/divyegala) URL: https://github.com/rapidsai/raft/pull/2432 --- cpp/include/raft/sparse/op/detail/sort.h | 10 +++++----- cpp/include/raft/sparse/op/sort.cuh | 14 +++++++------- 2 files changed, 12 insertions(+), 12 deletions(-) diff --git a/cpp/include/raft/sparse/op/detail/sort.h b/cpp/include/raft/sparse/op/detail/sort.h index 85ae825035..02287c2367 100644 --- a/cpp/include/raft/sparse/op/detail/sort.h +++ b/cpp/include/raft/sparse/op/detail/sort.h @@ -68,8 +68,8 @@ struct TupleComp { * @param vals vals array from coo matrix * @param stream: cuda stream to use */ -template -void coo_sort(int m, int n, int nnz, int* rows, int* cols, T* vals, cudaStream_t stream) +template +void coo_sort(IdxT m, IdxT n, IdxT nnz, IdxT* rows, IdxT* cols, T* vals, cudaStream_t stream) { auto coo_indices = thrust::make_zip_iterator(thrust::make_tuple(rows, cols)); @@ -83,10 +83,10 @@ void coo_sort(int m, int n, int nnz, int* rows, int* cols, T* vals, cudaStream_t * @param in: COO to sort by row * @param stream: the cuda stream to use */ -template -void coo_sort(COO* const in, cudaStream_t stream) +template +void coo_sort(COO* const in, cudaStream_t stream) { - coo_sort(in->n_rows, in->n_cols, in->nnz, in->rows(), in->cols(), in->vals(), stream); + coo_sort(in->n_rows, in->n_cols, in->nnz, in->rows(), in->cols(), in->vals(), stream); } /** diff --git a/cpp/include/raft/sparse/op/sort.cuh b/cpp/include/raft/sparse/op/sort.cuh index c6c3c2e220..5b8a792429 100644 --- a/cpp/include/raft/sparse/op/sort.cuh +++ b/cpp/include/raft/sparse/op/sort.cuh @@ -1,5 +1,5 @@ /* - * Copyright (c) 2019-2023, NVIDIA CORPORATION. + * Copyright (c) 2019-2024, NVIDIA CORPORATION. * * Licensed under the Apache License, Version 2.0 (the "License"); * you may not use this file except in compliance with the License. @@ -37,8 +37,8 @@ namespace op { * @param vals vals array from coo matrix * @param stream: cuda stream to use */ -template -void coo_sort(int m, int n, int nnz, int* rows, int* cols, T* vals, cudaStream_t stream) +template +void coo_sort(IdxT m, IdxT n, IdxT nnz, IdxT* rows, IdxT* cols, T* vals, cudaStream_t stream) { detail::coo_sort(m, n, nnz, rows, cols, vals, stream); } @@ -49,10 +49,10 @@ void coo_sort(int m, int n, int nnz, int* rows, int* cols, T* vals, cudaStream_t * @param in: COO to sort by row * @param stream: the cuda stream to use */ -template -void coo_sort(COO* const in, cudaStream_t stream) +template +void coo_sort(COO* const in, cudaStream_t stream) { - coo_sort(in->n_rows, in->n_cols, in->nnz, in->rows(), in->cols(), in->vals(), stream); + coo_sort(in->n_rows, in->n_cols, in->nnz, in->rows(), in->cols(), in->vals(), stream); } /** @@ -75,4 +75,4 @@ void coo_sort_by_weight( }; // end NAMESPACE sparse }; // end NAMESPACE raft -#endif \ No newline at end of file +#endif From e60cd7d063c3ab7830959d5aef465461f2daf99f Mon Sep 17 00:00:00 2001 From: Bradley Dice Date: Wed, 18 Sep 2024 03:05:44 -0500 Subject: [PATCH 3/3] Update to flake8 7.1.1. (#2435) We need to update flake8 to fix a false-positive that appears with older flake8 versions on Python 3.12. Authors: - Bradley Dice (https://github.com/bdice) Approvers: - James Lamb (https://github.com/jameslamb) URL: https://github.com/rapidsai/raft/pull/2435 --- .pre-commit-config.yaml | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/.pre-commit-config.yaml b/.pre-commit-config.yaml index 88be628e55..458d8b1b51 100644 --- a/.pre-commit-config.yaml +++ b/.pre-commit-config.yaml @@ -18,7 +18,7 @@ repos: # Explicitly specify the pyproject.toml at the repo root, not per-project. args: ["--config", "pyproject.toml"] - repo: https://github.com/PyCQA/flake8 - rev: 5.0.4 + rev: 7.1.1 hooks: - id: flake8 args: ["--config=.flake8"]