Skip to content

Commit

Permalink
Cleaning up cusparse_wrappers (#441)
Browse files Browse the repository at this point in the history
Closes #289

Authors:
  - Corey J. Nolet (https://github.com/cjnolet)

Approvers:
  - Jiaming Yuan (https://github.com/trivialfis)
  - Chuck Hastings (https://github.com/ChuckHastings)
  - Seunghwa Kang (https://github.com/seunghwak)

URL: #441
  • Loading branch information
cjnolet authored Feb 8, 2022
1 parent 6422bae commit 29718bd
Show file tree
Hide file tree
Showing 32 changed files with 566 additions and 219 deletions.
2 changes: 1 addition & 1 deletion cpp/doxygen/Doxyfile.in
Original file line number Diff line number Diff line change
Expand Up @@ -815,7 +815,7 @@ RECURSIVE = YES

EXCLUDE = @CMAKE_CURRENT_SOURCE_DIR@/include/raft/sparse/linalg/symmetrize.hpp \ # Contains device code
@CMAKE_CURRENT_SOURCE_DIR@/include/raft/sparse/csr.hpp \ # Contains device code
@CMAKE_CURRENT_SOURCE_DIR@/include/raft/sparse/cusparse_wrappers.h
@CMAKE_CURRENT_SOURCE_DIR@/include/raft/sparse/detail/cusparse_wrappers.h

# The EXCLUDE_SYMLINKS tag can be used to select whether or not files or
# directories that are symbolic links (a Unix file system feature) are excluded
Expand Down
2 changes: 1 addition & 1 deletion cpp/include/raft/handle.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -38,7 +38,7 @@
#include <raft/interruptible.hpp>
#include <raft/linalg/detail/cublas_wrappers.hpp>
#include <raft/linalg/detail/cusolver_wrappers.hpp>
#include <raft/sparse/cusparse_wrappers.h>
#include <raft/sparse/detail/cusparse_macros.h>
#include <rmm/cuda_stream_pool.hpp>
#include <rmm/exec_policy.hpp>

Expand Down
4 changes: 3 additions & 1 deletion cpp/include/raft/sparse/convert/dense.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -32,6 +32,7 @@ namespace convert {
* @param[in] handle : cusparse handle for conversion
* @param[in] nrows : number of rows in CSR
* @param[in] ncols : number of columns in CSR
* @param[in] nnz : number of nonzeros in CSR
* @param[in] csr_indptr : CSR row index pointer array
* @param[in] csr_indices : CSR column indices array
* @param[in] csr_data : CSR data array
Expand All @@ -44,6 +45,7 @@ template <typename value_idx, typename value_t>
void csr_to_dense(cusparseHandle_t handle,
value_idx nrows,
value_idx ncols,
value_idx nnz,
const value_idx* csr_indptr,
const value_idx* csr_indices,
const value_t* csr_data,
Expand All @@ -53,7 +55,7 @@ void csr_to_dense(cusparseHandle_t handle,
bool row_major = true)
{
detail::csr_to_dense<value_idx, value_t>(
handle, nrows, ncols, csr_indptr, csr_indices, csr_data, lda, out, stream, row_major);
handle, nrows, ncols, nnz, csr_indptr, csr_indices, csr_data, lda, out, stream, row_major);
}

}; // end NAMESPACE convert
Expand Down
2 changes: 1 addition & 1 deletion cpp/include/raft/sparse/convert/detail/coo.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -19,7 +19,7 @@
#include <cusparse_v2.h>
#include <raft/cuda_utils.cuh>
#include <raft/cudart_utils.h>
#include <raft/sparse/cusparse_wrappers.h>
#include <raft/sparse/detail/cusparse_wrappers.h>

#include <thrust/device_ptr.h>
#include <thrust/scan.h>
Expand Down
11 changes: 6 additions & 5 deletions cpp/include/raft/sparse/convert/detail/csr.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -21,7 +21,7 @@
#include <raft/cuda_utils.cuh>
#include <raft/cudart_utils.h>
#include <raft/handle.hpp>
#include <raft/sparse/cusparse_wrappers.h>
#include <raft/sparse/detail/cusparse_wrappers.h>
#include <rmm/device_uvector.hpp>

#include <thrust/device_ptr.h>
Expand Down Expand Up @@ -61,15 +61,16 @@ void coo_to_csr(const raft::handle_t& handle,
cudaMemcpyAsync(dstRows.data(), srcRows, sizeof(int) * nnz, cudaMemcpyDeviceToDevice, stream));
RAFT_CUDA_TRY(
cudaMemcpyAsync(dstCols, srcCols, sizeof(int) * nnz, cudaMemcpyDeviceToDevice, stream));
auto buffSize = raft::sparse::cusparsecoosort_bufferSizeExt(
auto buffSize = raft::sparse::detail::cusparsecoosort_bufferSizeExt(
cusparseHandle, m, m, nnz, srcRows, srcCols, stream);
rmm::device_uvector<char> pBuffer(buffSize, stream);
rmm::device_uvector<int> P(nnz, stream);
RAFT_CUSPARSE_TRY(cusparseCreateIdentityPermutation(cusparseHandle, nnz, P.data()));
raft::sparse::cusparsecoosortByRow(
raft::sparse::detail::cusparsecoosortByRow(
cusparseHandle, m, m, nnz, dstRows.data(), dstCols, P.data(), pBuffer.data(), stream);
raft::sparse::cusparsegthr(cusparseHandle, nnz, srcVals, dstVals, P.data(), stream);
raft::sparse::cusparsecoo2csr(cusparseHandle, dstRows.data(), nnz, m, dst_offsets, stream);
raft::sparse::detail::cusparsegthr(cusparseHandle, nnz, srcVals, dstVals, P.data(), stream);
raft::sparse::detail::cusparsecoo2csr(
cusparseHandle, dstRows.data(), nnz, m, dst_offsets, stream);
RAFT_CUDA_TRY(cudaDeviceSynchronize());
}

Expand Down
35 changes: 32 additions & 3 deletions cpp/include/raft/sparse/convert/detail/dense.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -19,7 +19,7 @@
#include <cusparse_v2.h>
#include <raft/cuda_utils.cuh>
#include <raft/cudart_utils.h>
#include <raft/sparse/cusparse_wrappers.h>
#include <raft/sparse/detail/cusparse_wrappers.h>

#include <thrust/device_ptr.h>
#include <thrust/scan.h>
Expand All @@ -31,6 +31,7 @@
#include <iostream>

#include <raft/sparse/detail/utils.h>
#include <rmm/device_uvector.hpp>

namespace raft {
namespace sparse {
Expand Down Expand Up @@ -67,6 +68,7 @@ __global__ void csr_to_dense_warp_per_row_kernel(
* @param[in] handle : cusparse handle for conversion
* @param[in] nrows : number of rows in CSR
* @param[in] ncols : number of columns in CSR
* @param[in] nnz : the number of nonzeros in CSR
* @param[in] csr_indptr : CSR row index pointer array
* @param[in] csr_indices : CSR column indices array
* @param[in] csr_data : CSR data array
Expand All @@ -79,6 +81,7 @@ template <typename value_idx, typename value_t>
void csr_to_dense(cusparseHandle_t handle,
value_idx nrows,
value_idx ncols,
value_idx nnz,
const value_idx* csr_indptr,
const value_idx* csr_indices,
const value_t* csr_data,
Expand All @@ -96,8 +99,34 @@ void csr_to_dense(cusparseHandle_t handle,
RAFT_CUSPARSE_TRY(cusparseSetMatIndexBase(out_mat, CUSPARSE_INDEX_BASE_ZERO));
RAFT_CUSPARSE_TRY(cusparseSetMatType(out_mat, CUSPARSE_MATRIX_TYPE_GENERAL));

RAFT_CUSPARSE_TRY(raft::sparse::cusparsecsr2dense(
handle, nrows, ncols, out_mat, csr_data, csr_indptr, csr_indices, out, lda, stream));
size_t buffer_size;
RAFT_CUSPARSE_TRY(raft::sparse::detail::cusparsecsr2dense_buffersize(handle,
nrows,
ncols,
nnz,
out_mat,
csr_data,
csr_indptr,
csr_indices,
out,
lda,
&buffer_size,
stream));

rmm::device_uvector<char> buffer(buffer_size, stream);

RAFT_CUSPARSE_TRY(raft::sparse::detail::cusparsecsr2dense(handle,
nrows,
ncols,
nnz,
out_mat,
csr_data,
csr_indptr,
csr_indices,
out,
lda,
buffer.data(),
stream));

RAFT_CUSPARSE_TRY_NO_THROW(cusparseDestroyMatDescr(out_mat));

Expand Down
2 changes: 1 addition & 1 deletion cpp/include/raft/sparse/detail/csr.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -19,7 +19,7 @@
#include <cusparse_v2.h>
#include <raft/cuda_utils.cuh>
#include <raft/cudart_utils.h>
#include <raft/sparse/cusparse_wrappers.h>
#include <raft/sparse/detail/cusparse_wrappers.h>
#include <rmm/device_scalar.hpp>
#include <rmm/device_uvector.hpp>

Expand Down
123 changes: 123 additions & 0 deletions cpp/include/raft/sparse/detail/cusparse_macros.h
Original file line number Diff line number Diff line change
@@ -0,0 +1,123 @@
/*
* Copyright (c) 2019-2021, NVIDIA CORPORATION.
*
* Licensed under the Apache License, Version 2.0 (the "License");
* you may not use this file except in compliance with the License.
* You may obtain a copy of the License at
*
* http://www.apache.org/licenses/LICENSE-2.0
*
* Unless required by applicable law or agreed to in writing, software
* distributed under the License is distributed on an "AS IS" BASIS,
* WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
* See the License for the specific language governing permissions and
* limitations under the License.
*/

#pragma once

#include <cusparse.h>
#include <raft/error.hpp>
///@todo: enable this once logging is enabled
//#include <cuml/common/logger.hpp>

#define _CUSPARSE_ERR_TO_STR(err) \
case err: return #err;

// Notes:
//(1.) CUDA_VER_10_1_UP aggregates all the CUDA version selection logic;
//(2.) to enforce a lower version,
//
//`#define CUDA_ENFORCE_LOWER
// #include <raft/sparse/detail/cusparse_wrappers.h>`
//
// (i.e., before including this header)
//
#define CUDA_VER_10_1_UP (CUDART_VERSION >= 10100)

namespace raft {

/**
* @brief Exception thrown when a cuSparse error is encountered.
*/
struct cusparse_error : public raft::exception {
explicit cusparse_error(char const* const message) : raft::exception(message) {}
explicit cusparse_error(std::string const& message) : raft::exception(message) {}
};

namespace sparse {
namespace detail {

inline const char* cusparse_error_to_string(cusparseStatus_t err)
{
#if defined(CUDART_VERSION) && CUDART_VERSION >= 10100
return cusparseGetErrorString(err);
#else // CUDART_VERSION
switch (err) {
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_SUCCESS);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_NOT_INITIALIZED);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_ALLOC_FAILED);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_INVALID_VALUE);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_ARCH_MISMATCH);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_EXECUTION_FAILED);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_INTERNAL_ERROR);
_CUSPARSE_ERR_TO_STR(CUSPARSE_STATUS_MATRIX_TYPE_NOT_SUPPORTED);
default: return "CUSPARSE_STATUS_UNKNOWN";
};
#endif // CUDART_VERSION
}

} // namespace detail
} // namespace sparse
} // namespace raft

#undef _CUSPARSE_ERR_TO_STR

/**
* @brief Error checking macro for cuSparse runtime API functions.
*
* Invokes a cuSparse runtime API function call, if the call does not return
* CUSPARSE_STATUS_SUCCESS, throws an exception detailing the cuSparse error that occurred
*/
#define RAFT_CUSPARSE_TRY(call) \
do { \
cusparseStatus_t const status = (call); \
if (CUSPARSE_STATUS_SUCCESS != status) { \
std::string msg{}; \
SET_ERROR_MSG(msg, \
"cuSparse error encountered at: ", \
"call='%s', Reason=%d:%s", \
#call, \
status, \
raft::sparse::detail::cusparse_error_to_string(status)); \
throw raft::cusparse_error(msg); \
} \
} while (0)

// FIXME: Remove after consumer rename
#ifndef CUSPARSE_TRY
#define CUSPARSE_TRY(call) RAFT_CUSPARSE_TRY(call)
#endif

// FIXME: Remove after consumer rename
#ifndef CUSPARSE_CHECK
#define CUSPARSE_CHECK(call) CUSPARSE_TRY(call)
#endif

//@todo: use logger here once logging is enabled
/** check for cusparse runtime API errors but do not assert */
#define RAFT_CUSPARSE_TRY_NO_THROW(call) \
do { \
cusparseStatus_t err = call; \
if (err != CUSPARSE_STATUS_SUCCESS) { \
printf("CUSPARSE call='%s' got errorcode=%d err=%s", \
#call, \
err, \
raft::sparse::detail::cusparse_error_to_string(err)); \
} \
} while (0)

// FIXME: Remove after consumer rename
#ifndef CUSPARSE_CHECK_NO_THROW
#define CUSPARSE_CHECK_NO_THROW(call) RAFT_CUSPARSE_TRY_NO_THROW(call)
#endif
Loading

0 comments on commit 29718bd

Please sign in to comment.