ViennaCL - The Vienna Computing Library  1.7.0
Free open-source GPU-accelerated linear algebra and solver library.
sparse_matrix_operations.hpp File Reference

Implementations of operations using sparse matrices using CUDA. More...

Go to the source code of this file.

Classes

struct  viennacl::linalg::cuda::mat_mult_matrix_index< LayoutT >
 Helper struct for accessing an element of a row- or column-major matrix. More...
 

Namespaces

 viennacl
 Main namespace in ViennaCL. Holds all the basic types such as vector, matrix, etc. and defines operations upon them.
 
 viennacl::linalg
 Provides all linear algebra operations which are not covered by operator overloads.
 
 viennacl::linalg::cuda
 Holds all CUDA compute kernels used by ViennaCL.
 
 viennacl::linalg::cuda::detail
 Helper functions for the CUDA linear algebra backend.
 

Functions

template<typename NumericT >
__global__ void viennacl::linalg::cuda::detail::csr_row_info_extractor_kernel (const unsigned int *row_indices, const unsigned int *column_indices, const NumericT *elements, NumericT *result, unsigned int size, unsigned int option)
 
template<typename NumericT , unsigned int AligmentV>
void viennacl::linalg::cuda::detail::row_info (compressed_matrix< NumericT, AligmentV > const &mat, vector_base< NumericT > &vec, viennacl::linalg::detail::row_info_types info_selector)
 
template<unsigned int SubWarpSizeV, typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_matrix_vec_mul_kernel (const unsigned int *row_indices, const unsigned int *column_indices, const NumericT *elements, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int size_result)
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_matrix_vec_mul_adaptive_kernel (const unsigned int *row_indices, const unsigned int *column_indices, const unsigned int *row_blocks, const NumericT *elements, unsigned int num_blocks, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int size_result)
 
template<class NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::compressed_matrix< NumericT, AlignmentV > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a compressed_matrix. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_matrix_d_mat_mul_kernel (const unsigned int *sp_mat_row_indices, const unsigned int *sp_mat_col_indices, const NumericT *sp_mat_elements, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::compressed_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_base< NumericT > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out sparse_matrix-dense_matrix multiplication first matrix being compressed. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_matrix_d_tr_mat_mul_kernel (const unsigned int *sp_mat_row_indices, const unsigned int *sp_mat_col_indices, const NumericT *sp_mat_elements, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::compressed_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_expression< const viennacl::matrix_base< NumericT >, const viennacl::matrix_base< NumericT >, viennacl::op_trans > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out matrix-trans(matrix) multiplication first matrix being compressed and the second transposed. More...
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_matrix_diagonal_kernel (const unsigned int *row_indices, const unsigned int *column_indices, const NumericT *elements, NumericT *result, unsigned int size)
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const SparseMatrixT &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::unit_lower_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const SparseMatrixT &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::lower_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const SparseMatrixT &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::unit_upper_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const SparseMatrixT &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::upper_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const matrix_expression< const SparseMatrixT, const SparseMatrixT, op_trans > &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::unit_lower_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const matrix_expression< const SparseMatrixT, const SparseMatrixT, op_trans > &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::lower_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const matrix_expression< const SparseMatrixT, const SparseMatrixT, op_trans > &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::unit_upper_tag)
 Carries out triangular inplace solves. More...
 
template<typename SparseMatrixT , typename NumericT >
viennacl::enable_if< viennacl::is_any_sparse_matrix< SparseMatrixT >::value >::type viennacl::linalg::cuda::inplace_solve (const matrix_expression< const SparseMatrixT, const SparseMatrixT, op_trans > &mat, viennacl::vector_base< NumericT > &vec, viennacl::linalg::upper_tag)
 Carries out triangular inplace solves. More...
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::detail::block_inplace_solve (const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &L, viennacl::backend::mem_handle const &block_indices, vcl_size_t num_blocks, vector_base< NumericT > const &, vector_base< NumericT > &vec, viennacl::linalg::unit_lower_tag)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::detail::block_inplace_solve (const matrix_expression< const compressed_matrix< NumericT, AlignmentV >, const compressed_matrix< NumericT, AlignmentV >, op_trans > &U, viennacl::backend::mem_handle const &block_indices, vcl_size_t num_blocks, vector_base< NumericT > const &U_diagonal, vector_base< NumericT > &vec, viennacl::linalg::upper_tag)
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::compressed_compressed_matrix_vec_mul_kernel (const unsigned int *row_jumper, const unsigned int *row_indices, const unsigned int *column_indices, const NumericT *elements, unsigned int nonzero_rows, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int size_result)
 
template<typename NumericT >
void viennacl::linalg::cuda::prod_impl (const viennacl::compressed_compressed_matrix< NumericT > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a compressed_compressed_matrix. More...
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::detail::coo_row_info_extractor (const unsigned int *coords, const NumericT *elements, const unsigned int *group_boundaries, NumericT *result, unsigned int option)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::detail::row_info (coordinate_matrix< NumericT, AlignmentV > const &mat, vector_base< NumericT > &vec, viennacl::linalg::detail::row_info_types info_selector)
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::coordinate_matrix_vec_mul_kernel (const unsigned int *coords, const NumericT *elements, const unsigned int *group_boundaries, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::coordinate_matrix< NumericT, AlignmentV > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a coordinate_matrix. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::coordinate_matrix_d_mat_mul_kernel (const unsigned int *coords, const NumericT *elements, const unsigned int *group_boundaries, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::coordinate_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_base< NumericT > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out Compressed Matrix(COO)-Dense Matrix multiplication. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::coordinate_matrix_d_tr_mat_mul_kernel (const unsigned int *coords, const NumericT *elements, const unsigned int *group_boundaries, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::coordinate_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_expression< const viennacl::matrix_base< NumericT >, const viennacl::matrix_base< NumericT >, viennacl::op_trans > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out Compressed Matrix(COO)-Dense Transposed Matrix multiplication. More...
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::ell_matrix_vec_mul_kernel (const unsigned int *coords, const NumericT *elements, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int row_num, unsigned int col_num, unsigned int internal_row_num, unsigned int items_per_row, unsigned int aligned_items_per_row)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::ell_matrix< NumericT, AlignmentV > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a ell_matrix. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::ell_matrix_d_mat_mul_kernel (const unsigned int *sp_mat_coords, const NumericT *sp_mat_elements, unsigned int sp_mat_row_num, unsigned int sp_mat_col_num, unsigned int sp_mat_internal_row_num, unsigned int sp_mat_items_per_row, unsigned int sp_mat_aligned_items_per_row, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::ell_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_base< NumericT > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out Sparse Matrix(ELL)-Dense Matrix multiplication. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::ell_matrix_d_tr_mat_mul_kernel (const unsigned int *sp_mat_coords, const NumericT *sp_mat_elements, unsigned int sp_mat_row_num, unsigned int sp_mat_col_num, unsigned int sp_mat_internal_row_num, unsigned int sp_mat_items_per_row, unsigned int sp_mat_aligned_items_per_row, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::ell_matrix< NumericT, AlignmentV > &sp_mat, const viennacl::matrix_expression< const viennacl::matrix_base< NumericT >, const viennacl::matrix_base< NumericT >, viennacl::op_trans > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out Sparse Matrix(ELL)-Dense Transposed Matrix multiplication. More...
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::sliced_ell_matrix_vec_mul_kernel (const unsigned int *columns_per_block, const unsigned int *column_indices, const unsigned int *block_start, const NumericT *elements, const NumericT *x, unsigned int start_x, unsigned int inc_x, unsigned int size_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int size_result, unsigned int block_size)
 
template<typename NumericT , typename IndexT >
void viennacl::linalg::cuda::prod_impl (const viennacl::sliced_ell_matrix< NumericT, IndexT > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a sliced_ell_matrix. More...
 
template<typename NumericT >
__global__ void viennacl::linalg::cuda::hyb_matrix_vec_mul_kernel (const unsigned int *ell_coords, const NumericT *ell_elements, const unsigned int *csr_rows, const unsigned int *csr_cols, const NumericT *csr_elements, const NumericT *x, unsigned int start_x, unsigned int inc_x, NumericT *result, unsigned int start_result, unsigned int inc_result, unsigned int row_num, unsigned int internal_row_num, unsigned int items_per_row, unsigned int aligned_items_per_row)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::hyb_matrix< NumericT, AlignmentV > &mat, const viennacl::vector_base< NumericT > &vec, viennacl::vector_base< NumericT > &result)
 Carries out matrix-vector multiplication with a hyb_matrix. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::hyb_matrix_d_mat_mul_kernel (const unsigned int *ell_coords, const NumericT *ell_elements, const unsigned int *csr_rows, const unsigned int *csr_cols, const NumericT *csr_elements, unsigned int row_num, unsigned int internal_row_num, unsigned int items_per_row, unsigned int aligned_items_per_row, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::hyb_matrix< NumericT, AlignmentV > &mat, const viennacl::matrix_base< NumericT > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out matrix-vector multiplication with a hyb_matrix. More...
 
template<typename DMatIndexT , typename ResultIndexT , typename NumericT >
__global__ void viennacl::linalg::cuda::hyb_matrix_d_tr_mat_mul_kernel (const unsigned int *ell_coords, const NumericT *ell_elements, const unsigned int *csr_rows, const unsigned int *csr_cols, const NumericT *csr_elements, unsigned int row_num, unsigned int internal_row_num, unsigned int items_per_row, unsigned int aligned_items_per_row, const NumericT *d_mat, unsigned int d_mat_row_start, unsigned int d_mat_col_start, unsigned int d_mat_row_inc, unsigned int d_mat_col_inc, unsigned int d_mat_row_size, unsigned int d_mat_col_size, unsigned int d_mat_internal_rows, unsigned int d_mat_internal_cols, NumericT *result, unsigned int result_row_start, unsigned int result_col_start, unsigned int result_row_inc, unsigned int result_col_inc, unsigned int result_row_size, unsigned int result_col_size, unsigned int result_internal_rows, unsigned int result_internal_cols)
 
template<typename NumericT , unsigned int AlignmentV>
void viennacl::linalg::cuda::prod_impl (const viennacl::hyb_matrix< NumericT, AlignmentV > &mat, const viennacl::matrix_expression< const viennacl::matrix_base< NumericT >, const viennacl::matrix_base< NumericT >, viennacl::op_trans > &d_mat, viennacl::matrix_base< NumericT > &result)
 Carries out matrix-vector multiplication with a hyb_matrix. More...
 

Detailed Description

Implementations of operations using sparse matrices using CUDA.

Definition in file sparse_matrix_operations.hpp.