Sparse level 2 functions

Sparse level 2 functions#

This module contains all sparse level 2 routines.

The sparse level 2 routines describe operations between a matrix in sparse format and a vector in dense format.

hipsparseXcsrmv()#

hipsparseStatus_t hipsparseScsrmv(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, const float *alpha, const hipsparseMatDescr_t descrA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const float *x, const float *beta, float *y)#

hipsparseStatus_t hipsparseDcsrmv(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, const double *alpha, const hipsparseMatDescr_t descrA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const double *x, const double *beta, double *y)#

hipsparseStatus_t hipsparseCcsrmv(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, const hipComplex *alpha, const hipsparseMatDescr_t descrA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipComplex *x, const hipComplex *beta, hipComplex *y)#

hipsparseStatus_t hipsparseZcsrmv(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipDoubleComplex *x, const hipDoubleComplex *beta, hipDoubleComplex *y)#

Sparse matrix vector multiplication using CSR storage format.

hipsparseXcsrmv multiplies the scalar \(\alpha\) with a sparse \(m \times n\) matrix, defined in CSR storage format, and the dense vector \(x\) and adds the result to the dense vector \(y\) that is multiplied by the scalar \(\beta\), such that

\[ y := \alpha \cdot op(A) \cdot x + \beta \cdot y, \]

with

\[\begin{split} op(A) = \left\{ \begin{array}{ll} A, & \text{if transA == HIPSPARSE_OPERATION_NON_TRANSPOSE} \\ A^T, & \text{if transA == HIPSPARSE_OPERATION_TRANSPOSE} \\ A^H, & \text{if transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE} \end{array} \right. \end{split}\]

for(i = 0; i < m; ++i)
{
    y[i] = beta * y[i];

    for(j = csrRowPtr[i]; j < csrRowPtr[i + 1]; ++j)
    {
        y[i] = y[i] + alpha * csrVal[j] * x[csrColInd[j]];
    }
}

Deprecated:: This function is deprecated when using the CUDA backend (CUDA 10.0+) and will be removed in CUDA 11.0. This deprecation does not apply to the ROCm backend.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the sparse CSR matrix. Must be non-negative.
n – [in] number of columns of the sparse CSR matrix. Must be non-negative.
nnz – [in] number of non-zero entries of the sparse CSR matrix. Must be non-negative.
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix. Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrSortedValA – [in] array of nnz elements of the sparse CSR matrix.
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix.
csrSortedColIndA – [in] array of nnz elements containing the column indices of the sparse CSR matrix.
x – [in] array of n elements ( \(op(A) == A\)) or m elements ( \(op(A) == A^T\) or \(op(A) == A^H\)).
beta – [in] scalar \(\beta\).
y – [inout] array of m elements ( \(op(A) == A\)) or n elements ( \(op(A) == A^T\) or \(op(A) == A^H\)).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descrA, alpha or beta is nullptr, m, n or nnz is negative, or csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, x or y is nullptr.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0 ) * ( 1.0 ) + beta * ( 4.0 ) = (  31.1 )
    //         ( 3.0  0.0  4.0 ) * ( 2.0 )          ( 5.0 ) = (  62.0 )
    //         ( 5.0  6.0  0.0 ) * ( 3.0 )          ( 6.0 ) = (  70.7 )
    //         ( 7.0  0.0  8.0 ) *                  ( 7.0 ) = ( 123.8 )

    const int m   = 4;
    const int n   = 3;
    const int nnz = 8;

    // CSR row pointers
    std::vector<int> hcsrRowPtr = {0, 2, 4, 6, 8};

    // CSR column indices
    std::vector<int> hcsrColInd = {0, 2, 0, 2, 0, 1, 0, 2};

    // CSR values
    std::vector<double> hcsrVal = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0};

    // Transposition of the matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Scalar alpha and beta
    double alpha = 3.7;
    double beta  = 1.3;

    // x and y
    std::vector<double> hx = {1.0, 2.0, 3.0};
    std::vector<double> hy = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Offload data to device
    int*    dcsrRowPtr;
    int*    dcsrColInd;
    double* dcsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dcsrVal, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtr, hcsrRowPtr.data(), sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColInd, hcsrColInd.data(), sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrVal, hcsrVal.data(), sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx.data(), sizeof(double) * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy.data(), sizeof(double) * m, hipMemcpyHostToDevice));

    // Call dcsrmv to perform y = alpha * A x + beta * y
    HIPSPARSE_CHECK(hipsparseDcsrmv(
        handle, trans, m, n, nnz, &alpha, descr, dcsrVal, dcsrRowPtr, dcsrColInd, dx, &beta, dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy.data(), dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    std::cout << "hy" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hy[i] << " ";
    }
    std::cout << std::endl;

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dcsrRowPtr));
    HIP_CHECK(hipFree(dcsrColInd));
    HIP_CHECK(hipFree(dcsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0 ) * ( 1.0 ) + beta * ( 4.0 ) = (  31.1 )
    //         ( 3.0  0.0  4.0 ) * ( 2.0 )          ( 5.0 ) = (  62.0 )
    //         ( 5.0  6.0  0.0 ) * ( 3.0 )          ( 6.0 ) = (  70.7 )
    //         ( 7.0  0.0  8.0 ) *                  ( 7.0 ) = ( 123.8 )

    const int m   = 4;
    const int n   = 3;
    const int nnz = 8;

    // CSR row pointers
    int hcsrRowPtr[] = {0, 2, 4, 6, 8};

    // CSR column indices
    int hcsrColInd[] = {0, 2, 0, 2, 0, 1, 0, 2};

    // CSR values
    double hcsrVal[] = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0};

    // Transposition of the matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Scalar alpha and beta
    double alpha = 3.7;
    double beta  = 1.3;

    // x and y
    double hx[] = {1.0, 2.0, 3.0};
    double hy[] = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Offload data to device
    int*    dcsrRowPtr;
    int*    dcsrColInd;
    double* dcsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dcsrVal, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrVal, hcsrVal, sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx, sizeof(double) * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy, sizeof(double) * m, hipMemcpyHostToDevice));

    // Call dcsrmv to perform y = alpha * A x + beta * y
    HIPSPARSE_CHECK(hipsparseDcsrmv(
        handle, trans, m, n, nnz, &alpha, descr, dcsrVal, dcsrRowPtr, dcsrColInd, dx, &beta, dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy, dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    printf("hy\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hy[i]);
    }
    printf("\n");

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dcsrRowPtr));
    HIP_CHECK(hipFree(dcsrColInd));
    HIP_CHECK(hipFree(dcsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

program example_fortran_csrmv
    use iso_c_binding
    use hipsparse

    implicit none

    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer :: hipMalloc
            type(c_ptr) :: ptr
            integer(c_size_t), value :: size
        end function hipMalloc

        function hipFree(ptr) &
                bind(c, name = 'hipFree')
            use iso_c_binding
            implicit none
            integer :: hipFree
            type(c_ptr), value :: ptr
        end function hipFree

        function hipMemcpy(dst, src, size, kind) &
                bind(c, name = 'hipMemcpy')
            use iso_c_binding
            implicit none
            integer :: hipMemcpy
            type(c_ptr), value :: dst
            type(c_ptr), intent(in), value :: src
            integer(c_size_t), value :: size
            integer(c_int), value :: kind
        end function hipMemcpy

        function hipDeviceSynchronize() &
                bind(c, name = 'hipDeviceSynchronize')
            use iso_c_binding
            implicit none
            integer :: hipDeviceSynchronize
        end function hipDeviceSynchronize

        function hipDeviceReset() &
                bind(c, name = 'hipDeviceReset')
            use iso_c_binding
            implicit none
            integer :: hipDeviceReset
        end function hipDeviceReset
    end interface

    integer, target :: h_csr_row_ptr(5), h_csr_col_ind(8)
    real(8), target :: h_csr_val(8), h_x(3), h_y(4)

    type(c_ptr) :: d_csr_row_ptr
    type(c_ptr) :: d_csr_col_ind
    type(c_ptr) :: d_csr_val
    type(c_ptr) :: d_x
    type(c_ptr) :: d_y

    integer :: i
    integer(c_int) :: m, n, nnz
    integer(c_int) :: trans

    real(c_double), target :: alpha, beta

    type(c_ptr) :: handle
    type(c_ptr) :: descr

    integer :: version

!   Input data
    m = 4
    n = 3
    nnz = 8

!   Fill CSR structure
    h_csr_row_ptr = (/0, 2, 4, 6, 8/)
    h_csr_col_ind = (/0, 2, 0, 2, 0, 1, 0, 2/)
    h_csr_val     = (/1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0/)

!   Transposition
    trans = HIPSPARSE_OPERATION_NON_TRANSPOSE

!   Scalar alpha and beta
    alpha = 3.7
    beta  = 1.3

!   x and y vectors
    h_x = (/1.0, 2.0, 3.0/)
    h_y = (/4.0, 5.0, 6.0, 7.0/)

!   Allocate device memory
    call HIP_CHECK(hipMalloc(d_csr_row_ptr, int(m + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_col_ind, int(nnz, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_val, int(nnz, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_x, int(n, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_y, int(m, c_size_t) * 8))

!   Copy host data to device
    call HIP_CHECK(hipMemcpy(d_csr_row_ptr, c_loc(h_csr_row_ptr), int(m + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_col_ind, c_loc(h_csr_col_ind), int(nnz, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_val, c_loc(h_csr_val), int(nnz, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_x, c_loc(h_x), int(n, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_y, c_loc(h_y), int(m, c_size_t) * 8, 1))

!   Create hipSPARSE handle
    call HIPSPARSE_CHECK(hipsparseCreate(handle))

!   Create matrix descriptor
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr))

!   Get hipSPARSE version
    call HIPSPARSE_CHECK(hipsparseGetVersion(handle, version))

!   Print version on screen
    write(*,fmt='(A,I0,A,I0,A,I0)') 'hipSPARSE version: ', version / 100000, '.', &
        mod(version / 100, 1000), '.', mod(version, 100)

!   Call dcsrmv to perform y = alpha * A * x + beta * y
    call HIPSPARSE_CHECK(hipsparseDcsrmv(handle, &
                                        trans, &
                                        m, &
                                        n, &
                                        nnz, &
                                        c_loc(alpha), &
                                        descr, &
                                        d_csr_val, &
                                        d_csr_row_ptr, &
                                        d_csr_col_ind, &
                                        d_x, &
                                        c_loc(beta), &
                                        d_y))

!   Print result
    call HIP_CHECK(hipMemcpy(c_loc(h_y), d_y, int(m, c_size_t) * 8, 2))

    write(*,fmt='(A)',advance='no') 'hy:'
    do i = 1, m
        write(*,fmt='(A,F0.2)',advance='no') ' ', h_y(i)
    end do
    write(*,*)

!   Clear hipSPARSE
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr))
    call HIPSPARSE_CHECK(hipsparseDestroy(handle))

!   Clear device memory
    call HIP_CHECK(hipFree(d_csr_row_ptr))
    call HIP_CHECK(hipFree(d_csr_col_ind))
    call HIP_CHECK(hipFree(d_csr_val))
    call HIP_CHECK(hipFree(d_x))
    call HIP_CHECK(hipFree(d_y))

end program example_fortran_csrmv

hipsparseXcsrsv2_zeroPivot()#

hipsparseStatus_t hipsparseXcsrsv2_zeroPivot(hipsparseHandle_t handle, csrsv2Info_t info, int *position)#

hipsparseXcsrsv2_zeroPivot returns HIPSPARSE_STATUS_ZERO_PIVOT, if either a structural or numerical zero has been found during hipsparseScsrsv2_solve(), hipsparseDcsrsv2_solve(), hipsparseCcsrsv2_solve() or hipsparseZcsrsv2_solve() computation. The first zero pivot \(j\) at \(A_{j,j}\) is stored in position, using same index base as the CSR matrix.

position can be in host or device memory. If no zero pivot has been found, position is set to -1 and HIPSPARSE_STATUS_SUCCESS is returned instead.

Deprecated:: This function is deprecated when using the CUDA backend (CUDA 11.0+) and will be removed in CUDA 12.0. This deprecation does not apply to the ROCm backend.

Note

hipsparseXcsrsv2_zeroPivot is a blocking function. It might influence performance negatively.

Parameters:

handle – [in] handle to the hipsparse library context queue.
info – [in] structure that holds the information collected during the analysis step.
position – [inout] pointer to zero pivot \(j\), can be in host or device memory.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, info or position is nullptr.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_ZERO_PIVOT – zero pivot has been found.

hipsparseXcsrsv2_bufferSize()#

hipsparseStatus_t hipsparseScsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDcsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCcsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZcsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseXcsrsv2_bufferSize returns the size of the temporary storage buffer in bytes that is required by hipsparseScsrsv2_analysis() and hipsparseXcsrsv2_solve(). The temporary storage buffer must be allocated by the user.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the sparse CSR matrix.
nnz – [in] number of non-zero entries of the sparse CSR matrix.
descrA – [in] descriptor of the sparse CSR matrix.
csrSortedValA – [in] array of nnz elements of the sparse CSR matrix.
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix.
csrSortedColIndA – [in] array of nnz elements containing the column indices of the sparse CSR matrix.
info – [out] structure that holds the information collected during the analysis step.
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXcsrsv2_analysis() and hipsparseXcsrsv2_solve().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, nnz, descrA, csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, info or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrsv2_bufferSizeExt()#

hipsparseStatus_t hipsparseScsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDcsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCcsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZcsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseXcsrsv2_bufferSizeExt returns the size of the temporary storage buffer in bytes that is required by hipsparseXcsrsv2_analysis() and hipsparseScsrsv2_solve(). The temporary storage buffer must be allocated by the user.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the sparse CSR matrix.
nnz – [in] number of non-zero entries of the sparse CSR matrix.
descrA – [in] descriptor of the sparse CSR matrix.
csrSortedValA – [in] array of nnz elements of the sparse CSR matrix.
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix.
csrSortedColIndA – [in] array of nnz elements containing the column indices of the sparse CSR matrix.
info – [out] structure that holds the information collected during the analysis step.
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXcsrsv2_analysis() and hipsparseXcsrsv2_solve().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, nnz, descrA, csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, info or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrsv2_analysis()#

hipsparseStatus_t hipsparseScsrsv2_analysis(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseDcsrsv2_analysis(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseCcsrsv2_analysis(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseZcsrsv2_analysis(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseXcsrsv2_analysis performs the analysis step for hipsparseXcsrsv2_solve(). It is expected that this function will be executed only once for a given matrix and particular operation type.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the sparse CSR matrix.
nnz – [in] number of non-zero entries of the sparse CSR matrix.
descrA – [in] descriptor of the sparse CSR matrix.
csrSortedValA – [in] array of nnz elements of the sparse CSR matrix.
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix.
csrSortedColIndA – [in] array of nnz elements containing the column indices of the sparse CSR matrix.
info – [out] structure that holds the information collected during the analysis step.
policy – [in] HIPSPARSE_SOLVE_POLICY_NO_LEVEL or HIPSPARSE_SOLVE_POLICY_USE_LEVEL.
pBuffer – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, nnz, descr, csrSortedRowPtrA, csrSortedColIndA, info or pBuffer is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrsv2_solve()#

hipsparseStatus_t hipsparseScsrsv2_solve(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const float *alpha, const hipsparseMatDescr_t descrA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, const float *f, float *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseDcsrsv2_solve(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const double *alpha, const hipsparseMatDescr_t descrA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, const double *f, double *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseCcsrsv2_solve(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipComplex *alpha, const hipsparseMatDescr_t descrA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, const hipComplex *f, hipComplex *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseZcsrsv2_solve(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int nnz, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, csrsv2Info_t info, const hipDoubleComplex *f, hipDoubleComplex *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

Sparse triangular solve using CSR storage format.

hipsparseXcsrsv2_solve solves a sparse triangular linear system of a sparse \(m \times m\) matrix, defined in CSR storage format, a dense solution vector \(y\) and the right-hand side \(x\) that is multiplied by \(\alpha\), such that

\[ op(A) \cdot y = \alpha \cdot x, \]

with

\[\begin{split} op(A) = \left\{ \begin{array}{ll} A, & \text{if trans == HIPSPARSE_OPERATION_NON_TRANSPOSE} \\ A^T, & \text{if trans == HIPSPARSE_OPERATION_TRANSPOSE} \\ A^H, & \text{if trans == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE} \end{array} \right. \end{split}\]

Performing the above operation requires three steps. First, the user calls hipsparseXcsrsv2_bufferSize() (or hipsparseXcsrsv2_bufferSizeExt()) which will determine the size of the required temporary storage buffer. The user then allocates this buffer and calls hipsparseXcsrsv2_analysis() which will perform analysis on the sparse matrix \(op(A)\). Finally, the user completes the computation by calling hipsparseXcsrsv2_solve. The buffer size, buffer allocation, and analysis only need to be called once for a given sparse matrix \(op(A)\) while the computation stage can be repeatedly used with different \(x\) and \(y\) vectors. Once all calls to hipsparseXcsrsv2_solve are complete, the temporary buffer can be deallocated.

Solving a triangular system involves division by the diagonal elements. This means that if the sparse matrix is missing the diagonal entry (referred to as a structural zero) or the diagonal entry is zero (referred to as a numerical zero) then a division by zero would occur. hipsparseXcsrsv2_solve tracks the location of the first zero pivot (either numerical or structural zero). The zero pivot status can be checked calling hipsparseXcsrsv2_zeroPivot(). If hipsparseXcsrsv2_zeroPivot() returns HIPSPARSE_STATUS_SUCCESS, then no zero pivot was found and therefore the matrix does not have a structural or numerical zero.

The user can specify that the sparse matrix should be interpreted as having ones on the diagonal by setting the diagonal type on the descriptor descrA to HIPSPARSE_DIAG_TYPE_UNIT using hipsparseSetMatDiagType. If hipsparseDiagType_t == HIPSPARSE_DIAG_TYPE_UNIT, no zero pivot will be reported, even if \(A_{j,j} = 0\) for some \(j\).

The sparse CSR matrix passed to hipsparseXcsrsv2_solve does not actually have to be a triangular matrix. Instead the triangular upper or lower part of the sparse matrix is solved based on hipsparseFillMode_t set on the descriptor descrA. If the fill mode is set to HIPSPARSE_FILL_MODE_LOWER, then the lower triangular matrix is solved. If the fill mode is set to HIPSPARSE_FILL_MODE_UPPER then the upper triangular matrix is solved.

Note

The sparse CSR matrix has to be sorted. This can be achieved by calling hipsparseXcsrsort().

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE and transA == HIPSPARSE_OPERATION_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the sparse CSR matrix.
nnz – [in] number of non-zero entries of the sparse CSR matrix.
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix.
csrSortedValA – [in] array of nnz elements of the sparse CSR matrix.
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix.
csrSortedColIndA – [in] array of nnz elements containing the column indices of the sparse CSR matrix.
info – [in] structure that holds the information collected during the analysis step.
f – [in] array of m elements, holding the right-hand side.
x – [out] array of m elements, holding the solution.
policy – [in] HIPSPARSE_SOLVE_POLICY_NO_LEVEL or HIPSPARSE_SOLVE_POLICY_USE_LEVEL.
pBuffer – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, nnz, descrA, alpha, csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, f or x is invalid.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0  0.0 ) * ( x_0 ) = ( 32.0 )
    //         ( 3.0  2.0  4.0  1.0 ) * ( x_1 ) = ( 14.7 )
    //         ( 5.0  6.0  1.0  3.0 ) * ( x_2 ) = ( 33.6 )
    //         ( 7.0  0.0  8.0  0.6 ) * ( x_3 ) = ( 10.0 )

    const int m   = 4;
    const int nnz = 13;

    // CSR row pointers
    std::vector<int> hcsrRowPtr = {0, 2, 6, 10, 13};

    // CSR column indices
    std::vector<int> hcsrColInd = {0, 2, 0, 1, 2, 3, 0, 1, 2, 3, 0, 2, 3};

    // CSR values
    std::vector<double> hcsrVal = {1.0, 2.0, 3.0, 2.0, 4.0, 1.0, 5.0, 6.0, 1.0, 3.0, 7.0, 8.0, 0.6};

    // Transposition of the matrix
    hipsparseOperation_t   trans  = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseSolvePolicy_t policy = HIPSPARSE_SOLVE_POLICY_USE_LEVEL;

    // Scalar alpha
    double alpha = 1.0;

    // f and x
    std::vector<double> hf = {32.0, 14.7, 33.6, 10.0};
    std::vector<double> hx(m);

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Set index base on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatIndexBase(descr, HIPSPARSE_INDEX_BASE_ZERO));

    // Set fill mode on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatFillMode(descr, HIPSPARSE_FILL_MODE_LOWER));

    // Set diag type on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatDiagType(descr, HIPSPARSE_DIAG_TYPE_UNIT));

    // Csrsv info
    csrsv2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateCsrsv2Info(&info));

    // Offload data to device
    int*    dcsrRowPtr;
    int*    dcsrColInd;
    double* dcsrVal;
    double* df;
    double* dx;

    HIP_CHECK(hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dcsrVal, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&df, sizeof(double) * m));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * m));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtr, hcsrRowPtr.data(), sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColInd, hcsrColInd.data(), sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrVal, hcsrVal.data(), sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(df, hf.data(), sizeof(double) * m, hipMemcpyHostToDevice));

    int bufferSize = 0;
    HIPSPARSE_CHECK(hipsparseDcsrsv2_bufferSize(
        handle, trans, m, nnz, descr, dcsrVal, dcsrRowPtr, dcsrColInd, info, &bufferSize));

    void* dbuffer = nullptr;
    HIP_CHECK(hipMalloc((void**)&dbuffer, bufferSize));

    HIPSPARSE_CHECK(hipsparseDcsrsv2_analysis(
        handle, trans, m, nnz, descr, dcsrVal, dcsrRowPtr, dcsrColInd, info, policy, dbuffer));

    // Call dcsrsv to perform alpha * A * x = f
    HIPSPARSE_CHECK(hipsparseDcsrsv2_solve(handle,
                                           trans,
                                           m,
                                           nnz,
                                           &alpha,
                                           descr,
                                           dcsrVal,
                                           dcsrRowPtr,
                                           dcsrColInd,
                                           info,
                                           df,
                                           dx,
                                           policy,
                                           dbuffer));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hx.data(), dx, sizeof(double) * m, hipMemcpyDeviceToHost));

    std::cout << "hx" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hx[i] << " ";
    }
    std::cout << std::endl;

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroyCsrsv2Info(info));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dcsrRowPtr));
    HIP_CHECK(hipFree(dcsrColInd));
    HIP_CHECK(hipFree(dcsrVal));
    HIP_CHECK(hipFree(df));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0  0.0 ) * ( x_0 ) = ( 32.0 )
    //         ( 3.0  2.0  4.0  1.0 ) * ( x_1 ) = ( 14.7 )
    //         ( 5.0  6.0  1.0  3.0 ) * ( x_2 ) = ( 33.6 )
    //         ( 7.0  0.0  8.0  0.6 ) * ( x_3 ) = ( 10.0 )

    const int m   = 4;
    const int nnz = 13;

    // CSR row pointers
    int hcsrRowPtr[] = {0, 2, 6, 10, 13};

    // CSR column indices
    int hcsrColInd[] = {0, 2, 0, 1, 2, 3, 0, 1, 2, 3, 0, 2, 3};

    // CSR values
    double hcsrVal[] = {1.0, 2.0, 3.0, 2.0, 4.0, 1.0, 5.0, 6.0, 1.0, 3.0, 7.0, 8.0, 0.6};

    // Transposition of the matrix
    hipsparseOperation_t   trans  = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseSolvePolicy_t policy = HIPSPARSE_SOLVE_POLICY_USE_LEVEL;

    // Scalar alpha
    double alpha = 1.0;

    // f and x
    double hf[] = {32.0, 14.7, 33.6, 10.0};
    double hx[m];

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Set index base on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatIndexBase(descr, HIPSPARSE_INDEX_BASE_ZERO));

    // Set fill mode on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatFillMode(descr, HIPSPARSE_FILL_MODE_LOWER));

    // Set diag type on descriptor
    HIPSPARSE_CHECK(hipsparseSetMatDiagType(descr, HIPSPARSE_DIAG_TYPE_UNIT));

    // Csrsv info
    csrsv2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateCsrsv2Info(&info));

    // Offload data to device
    int*    dcsrRowPtr;
    int*    dcsrColInd;
    double* dcsrVal;
    double* df;
    double* dx;

    HIP_CHECK(hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dcsrVal, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&df, sizeof(double) * m));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrVal, hcsrVal, sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(df, hf, sizeof(double) * m, hipMemcpyHostToDevice));

    int bufferSize = 0;
    HIPSPARSE_CHECK(hipsparseDcsrsv2_bufferSize(
        handle, trans, m, nnz, descr, dcsrVal, dcsrRowPtr, dcsrColInd, info, &bufferSize));

    void* dbuffer = NULL;
    HIP_CHECK(hipMalloc((void**)&dbuffer, bufferSize));

    HIPSPARSE_CHECK(hipsparseDcsrsv2_analysis(
        handle, trans, m, nnz, descr, dcsrVal, dcsrRowPtr, dcsrColInd, info, policy, dbuffer));

    // Call dcsrsv to perform alpha * A * x = f
    HIPSPARSE_CHECK(hipsparseDcsrsv2_solve(handle,
                                           trans,
                                           m,
                                           nnz,
                                           &alpha,
                                           descr,
                                           dcsrVal,
                                           dcsrRowPtr,
                                           dcsrColInd,
                                           info,
                                           df,
                                           dx,
                                           policy,
                                           dbuffer));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hx, dx, sizeof(double) * m, hipMemcpyDeviceToHost));

    printf("hx\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hx[i]);
    }
    printf("\n");

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroyCsrsv2Info(info));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dcsrRowPtr));
    HIP_CHECK(hipFree(dcsrColInd));
    HIP_CHECK(hipFree(dcsrVal));
    HIP_CHECK(hipFree(df));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

program example_fortran_csrsv2
    use iso_c_binding
    use hipsparse

    implicit none

    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer :: hipMalloc
            type(c_ptr) :: ptr
            integer(c_size_t), value :: size
        end function hipMalloc

        function hipFree(ptr) &
                bind(c, name = 'hipFree')
            use iso_c_binding
            implicit none
            integer :: hipFree
            type(c_ptr), value :: ptr
        end function hipFree

        function hipMemcpy(dst, src, size, kind) &
                bind(c, name = 'hipMemcpy')
            use iso_c_binding
            implicit none
            integer :: hipMemcpy
            type(c_ptr), value :: dst
            type(c_ptr), intent(in), value :: src
            integer(c_size_t), value :: size
            integer(c_int), value :: kind
        end function hipMemcpy
    end interface

    integer, target :: h_csr_row_ptr(5), h_csr_col_ind(13)
    real(8), target :: h_csr_val(13), h_f(4), h_x(4)

    type(c_ptr) :: d_csr_row_ptr, d_csr_col_ind, d_csr_val
    type(c_ptr) :: d_f, d_x, d_buffer

    integer :: i
    integer(c_int) :: m, nnz, trans, policy
    integer(c_int), target :: buffer_size

    real(c_double), target :: alpha

    type(c_ptr) :: handle, descr, info

    integer :: version

!   Input data
    m = 4
    nnz = 13

!   Fill CSR structure
    h_csr_row_ptr = (/0, 2, 6, 10, 13/)
    h_csr_col_ind = (/0, 2, 0, 1, 2, 3, 0, 1, 2, 3, 0, 2, 3/)
    h_csr_val     = (/1.0d0, 2.0d0, 3.0d0, 2.0d0, 4.0d0, 1.0d0, 5.0d0, 6.0d0, 1.0d0, 3.0d0, 7.0d0, 8.0d0, 0.6d0/)

!   Transposition and policy
    trans = HIPSPARSE_OPERATION_NON_TRANSPOSE
    policy = HIPSPARSE_SOLVE_POLICY_USE_LEVEL

!   Scalar alpha
    alpha = 1.0d0

!   f and x vectors
    h_f = (/32.0d0, 14.7d0, 33.6d0, 10.0d0/)
    h_x = 0.0d0

!   Allocate device memory
    call HIP_CHECK(hipMalloc(d_csr_row_ptr, int(m + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_col_ind, int(nnz, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_val, int(nnz, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_f, int(m, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_x, int(m, c_size_t) * 8))

!   Copy host data to device
    call HIP_CHECK(hipMemcpy(d_csr_row_ptr, c_loc(h_csr_row_ptr), int(m + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_col_ind, c_loc(h_csr_col_ind), int(nnz, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_val, c_loc(h_csr_val), int(nnz, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_f, c_loc(h_f), int(m, c_size_t) * 8, 1))

!   Create hipSPARSE handle
    call HIPSPARSE_CHECK(hipsparseCreate(handle))

!   Create matrix descriptor
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr))

!   Set matrix properties
    call HIPSPARSE_CHECK(hipsparseSetMatIndexBase(descr, HIPSPARSE_INDEX_BASE_ZERO))
    call HIPSPARSE_CHECK(hipsparseSetMatFillMode(descr, HIPSPARSE_FILL_MODE_LOWER))
    call HIPSPARSE_CHECK(hipsparseSetMatDiagType(descr, HIPSPARSE_DIAG_TYPE_UNIT))

!   Create csrsv2 info
    call HIPSPARSE_CHECK(hipsparseCreateCsrsv2Info(info))

!   Get hipSPARSE version
    call HIPSPARSE_CHECK(hipsparseGetVersion(handle, version))
    write(*,fmt='(A,I0,A,I0,A,I0)') 'hipSPARSE version: ', version / 100000, '.', &
        mod(version / 100, 1000), '.', mod(version, 100)

!   Get buffer size
    call HIPSPARSE_CHECK(hipsparseDcsrsv2_bufferSize(handle, &
                                                     trans, &
                                                     m, &
                                                     nnz, &
                                                     descr, &
                                                     d_csr_val, &
                                                     d_csr_row_ptr, &
                                                     d_csr_col_ind, &
                                                     info, &
                                                     c_loc(buffer_size)))

!   Allocate buffer
    call HIP_CHECK(hipMalloc(d_buffer, int(buffer_size, c_size_t)))

!   Analysis step
    call HIPSPARSE_CHECK(hipsparseDcsrsv2_analysis(handle, &
                                                   trans, &
                                                   m, &
                                                   nnz, &
                                                   descr, &
                                                   d_csr_val, &
                                                   d_csr_row_ptr, &
                                                   d_csr_col_ind, &
                                                   info, &
                                                   policy, &
                                                   d_buffer))

!   Call dcsrsv2_solve to perform alpha * A * x = f
    call HIPSPARSE_CHECK(hipsparseDcsrsv2_solve(handle, &
                                                trans, &
                                                m, &
                                                nnz, &
                                                c_loc(alpha), &
                                                descr, &
                                                d_csr_val, &
                                                d_csr_row_ptr, &
                                                d_csr_col_ind, &
                                                info, &
                                                d_f, &
                                                d_x, &
                                                policy, &
                                                d_buffer))

!   Copy result back to host
    call HIP_CHECK(hipMemcpy(c_loc(h_x), d_x, int(m, c_size_t) * 8, 2))

!   Print result
    write(*,fmt='(A)',advance='no') 'hx:'
    do i = 1, m
        write(*,fmt='(A,F0.2)',advance='no') ' ', h_x(i)
    end do
    write(*,*)

!   Clear hipSPARSE
    call HIPSPARSE_CHECK(hipsparseDestroyCsrsv2Info(info))
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr))
    call HIPSPARSE_CHECK(hipsparseDestroy(handle))

!   Clear device memory
    call HIP_CHECK(hipFree(d_csr_row_ptr))
    call HIP_CHECK(hipFree(d_csr_col_ind))
    call HIP_CHECK(hipFree(d_csr_val))
    call HIP_CHECK(hipFree(d_f))
    call HIP_CHECK(hipFree(d_x))
    call HIP_CHECK(hipFree(d_buffer))

end program example_fortran_csrsv2

hipsparseXhybmv()#

hipsparseStatus_t hipsparseShybmv(hipsparseHandle_t handle, hipsparseOperation_t transA, const float *alpha, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, const float *x, const float *beta, float *y)#

hipsparseStatus_t hipsparseDhybmv(hipsparseHandle_t handle, hipsparseOperation_t transA, const double *alpha, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, const double *x, const double *beta, double *y)#

hipsparseStatus_t hipsparseChybmv(hipsparseHandle_t handle, hipsparseOperation_t transA, const hipComplex *alpha, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, const hipComplex *x, const hipComplex *beta, hipComplex *y)#

hipsparseStatus_t hipsparseZhybmv(hipsparseHandle_t handle, hipsparseOperation_t transA, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, const hipDoubleComplex *x, const hipDoubleComplex *beta, hipDoubleComplex *y)#

Sparse matrix vector multiplication using HYB storage format.

hipsparseXhybmv multiplies the scalar \(\alpha\) with a sparse \(m \times n\) matrix, defined in HYB storage format, and the dense vector \(x\) and adds the result to the dense vector \(y\) that is multiplied by the scalar \(\beta\), such that

\[ y := \alpha \cdot op(A) \cdot x + \beta \cdot y, \]

with

\[ op(A) = \left\{ \begin{array}{ll} A, & \text{if transA == HIPSPARSE_OPERATION_NON_TRANSPOSE} \end{array} \right. \]

Deprecated:: This function is deprecated when using the CUDA backend (CUDA 10.0+) and will be removed in CUDA 11.0. This deprecation does not apply to the ROCm backend.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse HYB matrix. Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
hybA – [in] matrix in HYB storage format.
x – [in] array of n elements ( \(op(A) == A\)) or m elements ( \(op(A) == A^T\) or \(op(A) == A^H\)).
beta – [in] scalar \(\beta\).
y – [inout] array of m elements ( \(op(A) == A\)) or n elements ( \(op(A) == A^T\) or \(op(A) == A^H\)).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descrA, alpha, beta or hybA is nullptr, or x or y is nullptr.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_ALLOC_FAILED – the buffer could not be allocated.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA is not HIPSPARSE_OPERATION_NON_TRANSPOSE, or hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // A sparse matrix
    // 1 0 3 4
    // 0 0 5 1
    // 0 2 0 0
    // 4 0 0 8
    std::vector<int>    hAptr = {0, 3, 5, 6, 8};
    std::vector<int>    hAcol = {0, 2, 3, 2, 3, 1, 0, 3};
    std::vector<double> hAval = {1.0, 3.0, 4.0, 5.0, 1.0, 2.0, 4.0, 8.0};

    int m   = 4;
    int n   = 4;
    int nnz = 8;

    double halpha = 1.0;
    double hbeta  = 0.0;

    std::vector<double> hx = {1.0, 2.0, 3.0, 4.0};
    std::vector<double> hy = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    // Offload data to device
    int*    dAptr = NULL;
    int*    dAcol = NULL;
    double* dAval = NULL;
    double* dx    = NULL;
    double* dy    = NULL;

    HIP_CHECK(hipMalloc((void**)&dAptr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dAcol, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dAval, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dAptr, hAptr.data(), sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dAcol, hAcol.data(), sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dAval, hAval.data(), sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx.data(), sizeof(double) * n, hipMemcpyHostToDevice));

    // Convert CSR matrix to HYB format
    hipsparseHybMat_t hybA;
    HIPSPARSE_CHECK(hipsparseCreateHybMat(&hybA));

    HIPSPARSE_CHECK(hipsparseDcsr2hyb(
        handle, m, n, descrA, dAval, dAptr, dAcol, hybA, 0, HIPSPARSE_HYB_PARTITION_AUTO));

    // Clean up CSR structures
    HIP_CHECK(hipFree(dAptr));
    HIP_CHECK(hipFree(dAcol));
    HIP_CHECK(hipFree(dAval));

    // Call hipsparse hybmv
    HIPSPARSE_CHECK(hipsparseDhybmv(
        handle, HIPSPARSE_OPERATION_NON_TRANSPOSE, &halpha, descrA, hybA, dx, &hbeta, dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy.data(), dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    std::cout << "hy" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hy[i] << " ";
    }
    std::cout << std::endl;

    // Clear up on device
    HIPSPARSE_CHECK(hipsparseDestroyHybMat(hybA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // A sparse matrix
    // 1 0 3 4
    // 0 0 5 1
    // 0 2 0 0
    // 4 0 0 8
    int    hAptr[5] = {0, 3, 5, 6, 8};
    int    hAcol[8] = {0, 2, 3, 2, 3, 1, 0, 3};
    double hAval[]  = {1.0, 3.0, 4.0, 5.0, 1.0, 2.0, 4.0, 8.0};

    int m   = 4;
    int n   = 4;
    int nnz = 8;

    double halpha = 1.0;
    double hbeta  = 0.0;

    double hx[] = {1.0, 2.0, 3.0, 4.0};
    double hy[] = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    // Offload data to device
    int*    dAptr = NULL;
    int*    dAcol = NULL;
    double* dAval = NULL;
    double* dx    = NULL;
    double* dy    = NULL;

    HIP_CHECK(hipMalloc((void**)&dAptr, sizeof(int) * (m + 1)));
    HIP_CHECK(hipMalloc((void**)&dAcol, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dAval, sizeof(double) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dAptr, hAptr, sizeof(int) * (m + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dAcol, hAcol, sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dAval, hAval, sizeof(double) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx, sizeof(double) * n, hipMemcpyHostToDevice));

    // Convert CSR matrix to HYB format
    hipsparseHybMat_t hybA;
    HIPSPARSE_CHECK(hipsparseCreateHybMat(&hybA));

    HIPSPARSE_CHECK(hipsparseDcsr2hyb(
        handle, m, n, descrA, dAval, dAptr, dAcol, hybA, 0, HIPSPARSE_HYB_PARTITION_AUTO));

    // Clean up CSR structures
    HIP_CHECK(hipFree(dAptr));
    HIP_CHECK(hipFree(dAcol));
    HIP_CHECK(hipFree(dAval));

    // Call hipsparse hybmv
    HIPSPARSE_CHECK(hipsparseDhybmv(
        handle, HIPSPARSE_OPERATION_NON_TRANSPOSE, &halpha, descrA, hybA, dx, &hbeta, dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy, dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    printf("hy\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hy[i]);
    }
    printf("\n");

    // Clear up on device
    HIPSPARSE_CHECK(hipsparseDestroyHybMat(hybA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

program example_fortran_hybmv
    use iso_c_binding
    use hipsparse

    implicit none

    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer :: hipMalloc
            type(c_ptr) :: ptr
            integer(c_size_t), value :: size
        end function hipMalloc

        function hipFree(ptr) &
                bind(c, name = 'hipFree')
            use iso_c_binding
            implicit none
            integer :: hipFree
            type(c_ptr), value :: ptr
        end function hipFree

        function hipMemcpy(dst, src, size, kind) &
                bind(c, name = 'hipMemcpy')
            use iso_c_binding
            implicit none
            integer :: hipMemcpy
            type(c_ptr), value :: dst
            type(c_ptr), intent(in), value :: src
            integer(c_size_t), value :: size
            integer(c_int), value :: kind
        end function hipMemcpy
    end interface

    integer, target :: h_a_ptr(5), h_a_col(8)
    real(8), target :: h_a_val(8), h_x(4), h_y(4)

    type(c_ptr) :: d_a_ptr, d_a_col, d_a_val, d_x, d_y

    integer :: i
    integer(c_int) :: m, n, nnz

    real(c_double), target :: alpha, beta

    type(c_ptr) :: handle, descr_a, hyb_a

    integer :: version

!   Input data
    m = 4
    n = 4
    nnz = 8

!   Fill CSR structure (to be converted to HYB)
    h_a_ptr = (/0, 3, 5, 6, 8/)
    h_a_col = (/0, 2, 3, 2, 3, 1, 0, 3/)
    h_a_val = (/1.0d0, 3.0d0, 4.0d0, 5.0d0, 1.0d0, 2.0d0, 4.0d0, 8.0d0/)

!   Scalar alpha and beta
    alpha = 1.0d0
    beta  = 0.0d0

!   x and y vectors
    h_x = (/1.0d0, 2.0d0, 3.0d0, 4.0d0/)
    h_y = (/4.0d0, 5.0d0, 6.0d0, 7.0d0/)

!   Allocate device memory for CSR matrix
    call HIP_CHECK(hipMalloc(d_a_ptr, int(m + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_a_col, int(nnz, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_a_val, int(nnz, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_x, int(n, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_y, int(m, c_size_t) * 8))

!   Copy host data to device
    call HIP_CHECK(hipMemcpy(d_a_ptr, c_loc(h_a_ptr), int(m + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_a_col, c_loc(h_a_col), int(nnz, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_a_val, c_loc(h_a_val), int(nnz, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_x, c_loc(h_x), int(n, c_size_t) * 8, 1))

!   Create hipSPARSE handle
    call HIPSPARSE_CHECK(hipsparseCreate(handle))

!   Create matrix descriptor
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr_a))

!   Get hipSPARSE version
    call HIPSPARSE_CHECK(hipsparseGetVersion(handle, version))
    write(*,fmt='(A,I0,A,I0,A,I0)') 'hipSPARSE version: ', version / 100000, '.', &
        mod(version / 100, 1000), '.', mod(version, 100)

!   Create HYB matrix
    call HIPSPARSE_CHECK(hipsparseCreateHybMat(hyb_a))

!   Convert CSR to HYB format
    call HIPSPARSE_CHECK(hipsparseDcsr2hyb(handle, &
                                           m, &
                                           n, &
                                           descr_a, &
                                           d_a_val, &
                                           d_a_ptr, &
                                           d_a_col, &
                                           hyb_a, &
                                           0, &
                                           HIPSPARSE_HYB_PARTITION_AUTO))

!   Free CSR structures (no longer needed after conversion)
    call HIP_CHECK(hipFree(d_a_ptr))
    call HIP_CHECK(hipFree(d_a_col))
    call HIP_CHECK(hipFree(d_a_val))

!   Call dhybmv to perform y = alpha * A * x + beta * y
    call HIPSPARSE_CHECK(hipsparseDhybmv(handle, &
                                        HIPSPARSE_OPERATION_NON_TRANSPOSE, &
                                        c_loc(alpha), &
                                        descr_a, &
                                        hyb_a, &
                                        d_x, &
                                        c_loc(beta), &
                                        d_y))

!   Copy result back to host
    call HIP_CHECK(hipMemcpy(c_loc(h_y), d_y, int(m, c_size_t) * 8, 2))

!   Print result
    write(*,fmt='(A)',advance='no') 'hy:'
    do i = 1, m
        write(*,fmt='(A,F0.2)',advance='no') ' ', h_y(i)
    end do
    write(*,*)

!   Clear hipSPARSE
    call HIPSPARSE_CHECK(hipsparseDestroyHybMat(hyb_a))
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr_a))
    call HIPSPARSE_CHECK(hipsparseDestroy(handle))

!   Clear device memory
    call HIP_CHECK(hipFree(d_x))
    call HIP_CHECK(hipFree(d_y))

end program example_fortran_hybmv

hipsparseXbsrmv()#

hipsparseStatus_t hipsparseSbsrmv(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nb, int nnzb, const float *alpha, const hipsparseMatDescr_t descrA, const float *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, const float *x, const float *beta, float *y)#

hipsparseStatus_t hipsparseDbsrmv(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nb, int nnzb, const double *alpha, const hipsparseMatDescr_t descrA, const double *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, const double *x, const double *beta, double *y)#

hipsparseStatus_t hipsparseCbsrmv(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nb, int nnzb, const hipComplex *alpha, const hipsparseMatDescr_t descrA, const hipComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, const hipComplex *x, const hipComplex *beta, hipComplex *y)#

hipsparseStatus_t hipsparseZbsrmv(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nb, int nnzb, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, const hipDoubleComplex *x, const hipDoubleComplex *beta, hipDoubleComplex *y)#

Sparse matrix vector multiplication using BSR storage format.

hipsparseXbsrmv multiplies the scalar \(\alpha\) with a sparse \(m \times n\) matrix, defined in BSR storage format, and the dense vector \(x\) and adds the result to the dense vector \(y\) that is multiplied by the scalar \(\beta\), such that

\[ y := \alpha \cdot op(A) \cdot x + \beta \cdot y, \]

with

\[ op(A) = \left\{ \begin{array}{ll} A, & \text{if trans == HIPSPARSE_OPERATION_NON_TRANSPOSE} \end{array} \right. \]

and where \(m = mb \times blockDim\) and \(n= nb \times blockDim\).

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dirA – [in] matrix storage of BSR blocks.
transA – [in] matrix operation type.
mb – [in] number of block rows of the sparse BSR matrix. Must be non-negative.
nb – [in] number of block columns of the sparse BSR matrix. Must be non-negative.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix. Must be non-negative.
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse BSR matrix. Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
bsrSortedValA – [in] array of nnzb blocks of the sparse BSR matrix.
bsrSortedRowPtrA – [in] array of mb+1 elements that point to the start of every block row of the sparse BSR matrix.
bsrSortedColIndA – [in] array of nnzb elements containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix. Must be positive.
x – [in] array of nb*blockDim elements ( \(op(A) = A\)) or mb*blockDim elements ( \(op(A) = A^T\) or \(op(A) = A^H\)).
beta – [in] scalar \(\beta\).
y – [inout] array of mb*blockDim elements ( \(op(A) = A\)) or nb*blockDim elements ( \(op(A) = A^T\) or \(op(A) = A^H\)).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descrA, alpha or beta is nullptr, mb, nb or nnzb is negative, blockDim is less than or equal to zero, or bsrSortedValA, bsrSortedRowPtrA, bsrSortedColIndA, x or y is nullptr.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA is not HIPSPARSE_OPERATION_NON_TRANSPOSE, or hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0 ) * ( 1.0 ) + beta * ( 4.0 ) = (  31.1 )
    //         ( 3.0  0.0  4.0 ) * ( 2.0 )          ( 5.0 ) = (  62.0 )
    //         ( 5.0  6.0  0.0 ) * ( 3.0 )          ( 6.0 ) = (  70.7 )
    //         ( 7.0  0.0  8.0 ) *                  ( 7.0 ) = ( 123.8 )

    // BSR block dimension
    const int bsr_dim = 2;

    // Number of block rows and columns
    const int mb = 2;
    const int nb = 2;

    // Number of non-zero blocks
    const int nnzb = 4;

    // Number of rows and columns
    const int m = mb * bsr_dim;
    const int n = nb * bsr_dim;

    // BSR row pointers
    std::vector<int> hbsrRowPtr = {0, 2, 4};

    // BSR column indices
    std::vector<int> hbsrColInd = {0, 1, 0, 1};

    // BSR values
    std::vector<double> hbsrVal
        = {1.0, 3.0, 0.0, 0.0, 2.0, 4.0, 0.0, 0.0, 5.0, 7.0, 6.0, 0.0, 0.0, 8.0, 0.0, 0.0};

    // Block storage in column major
    hipsparseDirection_t dir = HIPSPARSE_DIRECTION_COLUMN;

    // Transposition of the matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Scalar alpha and beta
    double alpha = 3.7;
    double beta  = 1.3;

    // x and y
    std::vector<double> hx = {1.0, 2.0, 3.0, 0.0};
    std::vector<double> hy = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Offload data to device
    int*    dbsrRowPtr;
    int*    dbsrColInd;
    double* dbsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1)));
    HIP_CHECK(hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb));
    HIP_CHECK(hipMalloc((void**)&dbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(
        hipMemcpy(dbsrRowPtr, hbsrRowPtr.data(), sizeof(int) * (mb + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dbsrColInd, hbsrColInd.data(), sizeof(int) * nnzb, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(
        dbsrVal, hbsrVal.data(), sizeof(double) * nnzb * bsr_dim * bsr_dim, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx.data(), sizeof(double) * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy.data(), sizeof(double) * m, hipMemcpyHostToDevice));

    // Call dbsrmv to perform y = alpha * A x + beta * y
    HIPSPARSE_CHECK(hipsparseDbsrmv(handle,
                                    dir,
                                    trans,
                                    mb,
                                    nb,
                                    nnzb,
                                    &alpha,
                                    descr,
                                    dbsrVal,
                                    dbsrRowPtr,
                                    dbsrColInd,
                                    bsr_dim,
                                    dx,
                                    &beta,
                                    dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy.data(), dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    std::cout << "hy" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hy[i] << " ";
    }
    std::cout << std::endl;

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dbsrRowPtr));
    HIP_CHECK(hipFree(dbsrColInd));
    HIP_CHECK(hipFree(dbsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // alpha * ( 1.0  0.0  2.0 ) * ( 1.0 ) + beta * ( 4.0 ) = (  31.1 )
    //         ( 3.0  0.0  4.0 ) * ( 2.0 )          ( 5.0 ) = (  62.0 )
    //         ( 5.0  6.0  0.0 ) * ( 3.0 )          ( 6.0 ) = (  70.7 )
    //         ( 7.0  0.0  8.0 ) *                  ( 7.0 ) = ( 123.8 )

    // BSR block dimension
    const int bsr_dim = 2;

    // Number of block rows and columns
    const int mb = 2;
    const int nb = 2;

    // Number of non-zero blocks
    const int nnzb = 4;

    // Number of rows and columns
    const int m = mb * bsr_dim;
    const int n = nb * bsr_dim;

    // BSR row pointers
    int hbsrRowPtr[] = {0, 2, 4};

    // BSR column indices
    int hbsrColInd[] = {0, 1, 0, 1};

    // BSR values
    double hbsrVal[]
        = {1.0, 3.0, 0.0, 0.0, 2.0, 4.0, 0.0, 0.0, 5.0, 7.0, 6.0, 0.0, 0.0, 8.0, 0.0, 0.0};

    // Block storage in column major
    hipsparseDirection_t dir = HIPSPARSE_DIRECTION_COLUMN;

    // Transposition of the matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Scalar alpha and beta
    double alpha = 3.7;
    double beta  = 1.3;

    // x and y
    double hx[] = {1.0, 2.0, 3.0, 0.0};
    double hy[] = {4.0, 5.0, 6.0, 7.0};

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Offload data to device
    int*    dbsrRowPtr;
    int*    dbsrColInd;
    double* dbsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1)));
    HIP_CHECK(hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb));
    HIP_CHECK(hipMalloc((void**)&dbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * n));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dbsrRowPtr, hbsrRowPtr, sizeof(int) * (mb + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dbsrColInd, hbsrColInd, sizeof(int) * nnzb, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(
        dbsrVal, hbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx, sizeof(double) * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy, sizeof(double) * m, hipMemcpyHostToDevice));

    // Call dbsrmv to perform y = alpha * A x + beta * y
    HIPSPARSE_CHECK(hipsparseDbsrmv(handle,
                                    dir,
                                    trans,
                                    mb,
                                    nb,
                                    nnzb,
                                    &alpha,
                                    descr,
                                    dbsrVal,
                                    dbsrRowPtr,
                                    dbsrColInd,
                                    bsr_dim,
                                    dx,
                                    &beta,
                                    dy));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy, dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    printf("hy\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hy[i]);
    }
    printf("\n");

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dbsrRowPtr));
    HIP_CHECK(hipFree(dbsrColInd));
    HIP_CHECK(hipFree(dbsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));

    return 0;
}

program example_fortran_bsrmv
    use iso_c_binding
    use hipsparse

    implicit none

    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer :: hipMalloc
            type(c_ptr) :: ptr
            integer(c_size_t), value :: size
        end function hipMalloc

        function hipFree(ptr) &
                bind(c, name = 'hipFree')
            use iso_c_binding
            implicit none
            integer :: hipFree
            type(c_ptr), value :: ptr
        end function hipFree

        function hipMemcpy(dst, src, size, kind) &
                bind(c, name = 'hipMemcpy')
            use iso_c_binding
            implicit none
            integer :: hipMemcpy
            type(c_ptr), value :: dst
            type(c_ptr), intent(in), value :: src
            integer(c_size_t), value :: size
            integer(c_int), value :: kind
        end function hipMemcpy
    end interface

    integer, target :: h_bsr_row_ptr(3), h_bsr_col_ind(4)
    real(8), target :: h_bsr_val(16), h_x(4), h_y(4)

    type(c_ptr) :: d_bsr_row_ptr, d_bsr_col_ind, d_bsr_val, d_x, d_y

    integer :: i
    integer(c_int) :: bsr_dim, mb, nb, nnzb, m, n, dir, trans

    real(c_double), target :: alpha, beta

    type(c_ptr) :: handle, descr

    integer :: version

!   BSR block dimension
    bsr_dim = 2

!   Number of block rows and columns
    mb = 2
    nb = 2

!   Number of non-zero blocks
    nnzb = 4

!   Number of rows and columns
    m = mb * bsr_dim
    n = nb * bsr_dim

!   Fill BSR structure
    h_bsr_row_ptr = (/0, 2, 4/)
    h_bsr_col_ind = (/0, 1, 0, 1/)
    h_bsr_val     = (/1.0d0, 3.0d0, 0.0d0, 0.0d0, 2.0d0, 4.0d0, 0.0d0, 0.0d0, &
                      5.0d0, 7.0d0, 6.0d0, 0.0d0, 0.0d0, 8.0d0, 0.0d0, 0.0d0/)

!   Block storage direction and transposition
    dir = HIPSPARSE_DIRECTION_COLUMN
    trans = HIPSPARSE_OPERATION_NON_TRANSPOSE

!   Scalar alpha and beta
    alpha = 3.7d0
    beta  = 1.3d0

!   x and y vectors
    h_x = (/1.0d0, 2.0d0, 3.0d0, 0.0d0/)
    h_y = (/4.0d0, 5.0d0, 6.0d0, 7.0d0/)

!   Allocate device memory
    call HIP_CHECK(hipMalloc(d_bsr_row_ptr, int(mb + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_bsr_col_ind, int(nnzb, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_bsr_val, int(nnzb * bsr_dim * bsr_dim, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_x, int(n, c_size_t) * 8))
    call HIP_CHECK(hipMalloc(d_y, int(m, c_size_t) * 8))

!   Copy host data to device
    call HIP_CHECK(hipMemcpy(d_bsr_row_ptr, c_loc(h_bsr_row_ptr), int(mb + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_bsr_col_ind, c_loc(h_bsr_col_ind), int(nnzb, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_bsr_val, c_loc(h_bsr_val), int(nnzb * bsr_dim * bsr_dim, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_x, c_loc(h_x), int(n, c_size_t) * 8, 1))
    call HIP_CHECK(hipMemcpy(d_y, c_loc(h_y), int(m, c_size_t) * 8, 1))

!   Create hipSPARSE handle
    call HIPSPARSE_CHECK(hipsparseCreate(handle))

!   Create matrix descriptor
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr))

!   Get hipSPARSE version
    call HIPSPARSE_CHECK(hipsparseGetVersion(handle, version))
    write(*,fmt='(A,I0,A,I0,A,I0)') 'hipSPARSE version: ', version / 100000, '.', &
        mod(version / 100, 1000), '.', mod(version, 100)

!   Call dbsrmv to perform y = alpha * A * x + beta * y
    call HIPSPARSE_CHECK(hipsparseDbsrmv(handle, &
                                        dir, &
                                        trans, &
                                        mb, &
                                        nb, &
                                        nnzb, &
                                        c_loc(alpha), &
                                        descr, &
                                        d_bsr_val, &
                                        d_bsr_row_ptr, &
                                        d_bsr_col_ind, &
                                        bsr_dim, &
                                        d_x, &
                                        c_loc(beta), &
                                        d_y))

!   Copy result back to host
    call HIP_CHECK(hipMemcpy(c_loc(h_y), d_y, int(m, c_size_t) * 8, 2))

!   Print result
    write(*,fmt='(A)',advance='no') 'hy:'
    do i = 1, m
        write(*,fmt='(A,F0.2)',advance='no') ' ', h_y(i)
    end do
    write(*,*)

!   Clear hipSPARSE
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr))
    call HIPSPARSE_CHECK(hipsparseDestroy(handle))

!   Clear device memory
    call HIP_CHECK(hipFree(d_bsr_row_ptr))
    call HIP_CHECK(hipFree(d_bsr_col_ind))
    call HIP_CHECK(hipFree(d_bsr_val))
    call HIP_CHECK(hipFree(d_x))
    call HIP_CHECK(hipFree(d_y))

end program example_fortran_bsrmv

hipsparseXbsrxmv()#

hipsparseStatus_t hipsparseSbsrxmv(hipsparseHandle_t handle, hipsparseDirection_t dir, hipsparseOperation_t trans, int sizeOfMask, int mb, int nb, int nnzb, const float *alpha, const hipsparseMatDescr_t descr, const float *bsrVal, const int *bsrMaskPtr, const int *bsrRowPtr, const int *bsrEndPtr, const int *bsrColInd, int blockDim, const float *x, const float *beta, float *y)#

hipsparseStatus_t hipsparseDbsrxmv(hipsparseHandle_t handle, hipsparseDirection_t dir, hipsparseOperation_t trans, int sizeOfMask, int mb, int nb, int nnzb, const double *alpha, const hipsparseMatDescr_t descr, const double *bsrVal, const int *bsrMaskPtr, const int *bsrRowPtr, const int *bsrEndPtr, const int *bsrColInd, int blockDim, const double *x, const double *beta, double *y)#

hipsparseStatus_t hipsparseCbsrxmv(hipsparseHandle_t handle, hipsparseDirection_t dir, hipsparseOperation_t trans, int sizeOfMask, int mb, int nb, int nnzb, const hipComplex *alpha, const hipsparseMatDescr_t descr, const hipComplex *bsrVal, const int *bsrMaskPtr, const int *bsrRowPtr, const int *bsrEndPtr, const int *bsrColInd, int blockDim, const hipComplex *x, const hipComplex *beta, hipComplex *y)#

hipsparseStatus_t hipsparseZbsrxmv(hipsparseHandle_t handle, hipsparseDirection_t dir, hipsparseOperation_t trans, int sizeOfMask, int mb, int nb, int nnzb, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descr, const hipDoubleComplex *bsrVal, const int *bsrMaskPtr, const int *bsrRowPtr, const int *bsrEndPtr, const int *bsrColInd, int blockDim, const hipDoubleComplex *x, const hipDoubleComplex *beta, hipDoubleComplex *y)#

Sparse matrix vector multiplication with mask operation using BSR storage format.

hipsparseXbsrxmv multiplies the scalar \(\alpha\) with a sparse \((mb \times \text{blockDim}) \times (nb \times \text{blockDim})\) modified matrix, defined in BSR storage format, and the dense vector \(x\) and adds the result to the dense vector \(y\) that is multiplied by the scalar \(\beta\), such that

\[ y := \left( \alpha \cdot op(A) \cdot x + \beta \cdot y \right)\left( \text{mask} \right), \]

with

\[\begin{split} op(A) = \left\{ \begin{array}{ll} A, & \text{if trans == HIPSPARSE_OPERATION_NON_TRANSPOSE} \\ A^T, & \text{if trans == HIPSPARSE_OPERATION_TRANSPOSE} \\ A^H, & \text{if trans == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE} \end{array} \right. \end{split}\]

The \(\text{mask}\) is defined as an array of block row indices. The input sparse matrix is defined with a modified BSR storage format where the beginning and the end of each row is defined with two arrays, bsrRowPtr and bsr_end_ptr (both of size mb), rather the usual bsrRowPtr of size mb+1.

Deprecated:: This function is deprecated when using the CUDA backend (CUDA 12.0+) and will be removed in CUDA 13.0. This deprecation does not apply to the ROCm backend.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only trans == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported. Currently, blockDim == 1 is not supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dir – [in] matrix storage of BSR blocks.
trans – [in] matrix operation type.
sizeOfMask – [in] number of updated block rows of the array y. Must be non-negative and not greater than mb.
mb – [in] number of block rows of the sparse BSR matrix. Must be non-negative.
nb – [in] number of block columns of the sparse BSR matrix. Must be non-negative.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix. Must be non-negative.
alpha – [in] scalar \(\alpha\).
descr – [in] descriptor of the sparse BSR matrix. Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
bsrVal – [in] array of nnzb blocks of the sparse BSR matrix.
bsrMaskPtr – [in] array of sizeOfMask elements that give the indices of the updated block rows.
bsrRowPtr – [in] array of mb elements that point to the start of every block row of the sparse BSR matrix.
bsrEndPtr – [in] array of mb elements that point to the end of every block row of the sparse BSR matrix.
bsrColInd – [in] array of nnzb elements containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix. Must be greater than 1.
x – [in] array of nb*blockDim elements ( \(op(A) = A\)) or mb*blockDim elements ( \(op(A) = A^T\) or \(op(A) = A^H\)).
beta – [in] scalar \(\beta\).
y – [inout] array of mb*blockDim elements ( \(op(A) = A\)) or nb*blockDim elements ( \(op(A) = A^T\) or \(op(A) = A^H\)).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descr, alpha or beta is nullptr, mb, nb, nnzb or sizeOfMask is negative, sizeOfMask is greater than mb, blockDim is less than or equal to 1, or bsrVal, bsrMaskPtr, bsrRowPtr, bsrEndPtr, bsrColInd, x or y is nullptr.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_NOT_SUPPORTED – trans is not HIPSPARSE_OPERATION_NON_TRANSPOSE, or hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXbsrsv2_zeroPivot()#

hipsparseStatus_t hipsparseXbsrsv2_zeroPivot(hipsparseHandle_t handle, bsrsv2Info_t info, int *position)#

hipsparseXbsrsv2_zeroPivot returns HIPSPARSE_STATUS_ZERO_PIVOT, if either a structural or numerical zero has been found during hipsparseXbsrsv2_analysis() or hipsparseXbsrsv2_solve() computation. The first zero pivot \(j\) at \(A_{j,j}\) is stored in position, using same index base as the BSR matrix.

position can be in host or device memory. If no zero pivot has been found, position is set to -1 and HIPSPARSE_STATUS_SUCCESS is returned instead.

Deprecated:: This function is deprecated when using the CUDA backend (CUDA 12.0+) and will be removed in CUDA 13.0. This deprecation does not apply to the ROCm backend.

Note

hipsparseXbsrsv2_zeroPivot is a blocking function. It might influence performance negatively.

Parameters:

handle – [in] handle to the hipsparse library context queue.
info – [in] structure that holds the information collected during the analysis step.
position – [inout] pointer to zero pivot \(j\), can be in host or device memory.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, info or position is nullptr.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_ZERO_PIVOT – zero pivot has been found.

hipsparseXbsrsv2_bufferSize()#

hipsparseStatus_t hipsparseSbsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, float *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDbsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, double *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCbsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, hipComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZbsrsv2_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, hipDoubleComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, int *pBufferSizeInBytes)#

hipsparseXbsrsv2_bufferSize returns the size of the temporary storage buffer in bytes that is required by hipsparseXbsrsv2_analysis() and hipsparseXbsrsv2_solve(). The temporary storage buffer must be allocated by the user.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dirA – [in] matrix storage of BSR blocks.
transA – [in] matrix operation type.
mb – [in] number of block rows of the sparse BSR matrix.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix.
descrA – [in] descriptor of the sparse BSR matrix.
bsrSortedValA – [in] array of nnzb blocks of the sparse BSR matrix.
bsrSortedRowPtrA – [in] array of mb+1 elements that point to the start of every block row of the sparse BSR matrix.
bsrSortedColIndA – [in] array of nnz containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix.
info – [out] structure that holds the information collected during the analysis step.
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXbsrsv2_analysis() and hipsparseXbsrsv2_solve().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, mb, nnzb or blockDim, descr, bsrSortedValA, bsrSortedRowPtrA, bsrSortedColIndA, info or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXbsrsv2_bufferSizeExt()#

hipsparseStatus_t hipsparseSbsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, float *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDbsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, double *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCbsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, hipComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZbsrsv2_bufferSizeExt(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, hipDoubleComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseXbsrsv2_bufferSizeExt returns the size of the temporary storage buffer in bytes that is required by hipsparseXbsrsv2_analysis() and hipsparseXbsrsv2_solve(). The temporary storage buffer must be allocated by the user.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dirA – [in] matrix storage of BSR blocks.
transA – [in] matrix operation type.
mb – [in] number of block rows of the sparse BSR matrix.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix.
descrA – [in] descriptor of the sparse BSR matrix.
bsrSortedValA – [in] array of nnzb blocks of the sparse BSR matrix.
bsrSortedRowPtrA – [in] array of mb+1 elements that point to the start of every block row of the sparse BSR matrix.
bsrSortedColIndA – [in] array of nnz containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix.
info – [out] structure that holds the information collected during the analysis step.
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXbsrsv2_analysis() and hipsparseXbsrsv2_solve().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, mb, nnzb or blockDim, descr, bsrSortedValA, bsrSortedRowPtrA, bsrSortedColIndA, info or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXbsrsv2_analysis()#

hipsparseStatus_t hipsparseSbsrsv2_analysis(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, const float *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseDbsrsv2_analysis(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, const double *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseCbsrsv2_analysis(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, const hipComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseZbsrsv2_analysis(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseXbsrsv2_analysis performs the analysis step for hipsparseXbsrsv2_solve(). It is expected that this function will be executed only once for a given matrix and particular operation type.

Note

If the matrix sparsity pattern changes, the gathered information will become invalid.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dirA – [in] matrix storage of BSR blocks.
transA – [in] matrix operation type.
mb – [in] number of block rows of the sparse BSR matrix.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix.
descrA – [in] descriptor of the sparse BSR matrix.
bsrSortedValA – [in] array of nnzb blocks of the sparse BSR matrix.
bsrSortedRowPtrA – [in] array of mb+1 elements that point to the start of every block row of the sparse BSR matrix.
bsrSortedColIndA – [in] array of nnz containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix.
info – [out] structure that holds the information collected during the analysis step.
policy – [in] HIPSPARSE_SOLVE_POLICY_NO_LEVEL or HIPSPARSE_SOLVE_POLICY_USE_LEVEL.
pBuffer – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, mb, nnzb, blockDim, descrA, bsrSortedRowPtrA, bsrSortedColIndA, info or pBuffer is invalid.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXbsrsv2_solve()#

hipsparseStatus_t hipsparseSbsrsv2_solve(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const float *alpha, const hipsparseMatDescr_t descrA, const float *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, const float *f, float *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseDbsrsv2_solve(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const double *alpha, const hipsparseMatDescr_t descrA, const double *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, const double *f, double *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseCbsrsv2_solve(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipComplex *alpha, const hipsparseMatDescr_t descrA, const hipComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, const hipComplex *f, hipComplex *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

hipsparseStatus_t hipsparseZbsrsv2_solve(hipsparseHandle_t handle, hipsparseDirection_t dirA, hipsparseOperation_t transA, int mb, int nnzb, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrSortedValA, const int *bsrSortedRowPtrA, const int *bsrSortedColIndA, int blockDim, bsrsv2Info_t info, const hipDoubleComplex *f, hipDoubleComplex *x, hipsparseSolvePolicy_t policy, void *pBuffer)#

Sparse triangular solve using BSR storage format.

hipsparseXbsrsv2_solve solves a sparse triangular linear system of a sparse \(m \times m\) matrix, defined in BSR storage format, a dense solution vector \(y\) and the right-hand side \(x\) that is multiplied by \(\alpha\), such that

\[ op(A) \cdot y = \alpha \cdot x, \]

with

\[\begin{split} op(A) = \left\{ \begin{array}{ll} A, & \text{if trans == HIPSPARSE_OPERATION_NON_TRANSPOSE} \\ A^T, & \text{if trans == HIPSPARSE_OPERATION_TRANSPOSE} \end{array} \right. \end{split}\]

Performing the above operation requires three steps. First, the user calls hipsparseXbsrsv2_bufferSize() which will determine the size of the required temporary storage buffer. The user then allocates this buffer and calls hipsparseXbsrsv2_analysis() which will perform analysis on the sparse matrix \(op(A)\). Finally, the user completes the computation by calling hipsparseXbsrsv2_solve. The buffer size, buffer allocation, and analysis only need to be called once for a given sparse matrix \(op(A)\) while the computation stage can be repeatedly used with different \(x\) and \(y\) vectors. Once all calls to hipsparseXbsrsv2_solve are complete, the temporary buffer can be deallocated.

Solving a triangular system involves inverting the diagonal blocks. This means that if the sparse matrix is missing the diagonal block (referred to as a structural zero) or the diagonal block is not invertible (referred to as a numerical zero) then a solution is not possible. hipsparseXbsrsv2_solve tracks the location of the first zero pivot (either numerical or structural zero). The zero pivot status can be checked calling hipsparseXbsrsv2_zeroPivot(). If hipsparseXbsrsv2_zeroPivot() returns HIPSPARSE_STATUS_SUCCESS, then no zero pivot was found and therefore the matrix does not have a structural or numerical zero.

The user can specify that the sparse matrix should be interpreted as having identity blocks on the diagonal by setting the diagonal type on the descriptor descrA to HIPSPARSE_DIAG_TYPE_UNIT using hipsparseSetMatDiagType. If hipsparseDiagType_t == HIPSPARSE_DIAG_TYPE_UNIT, no zero pivot will be reported, even if the diagonal block \(A_{j,j}\) for some \(j\) is not invertible.

The sparse CSR matrix passed to hipsparseXbsrsv2_solve does not actually have to be a triangular matrix. Instead the triangular upper or lower part of the sparse matrix is solved based on hipsparseFillMode_t set on the descriptor descrA. If the fill mode is set to HIPSPARSE_FILL_MODE_LOWER, then the lower triangular matrix is solved. If the fill mode is set to HIPSPARSE_FILL_MODE_UPPER then the upper triangular matrix is solved.

Note

The sparse BSR matrix has to be sorted.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE and transA == HIPSPARSE_OPERATION_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
dirA – [in] matrix storage of BSR blocks.
transA – [in] matrix operation type.
mb – [in] number of block rows of the sparse BSR matrix.
nnzb – [in] number of non-zero blocks of the sparse BSR matrix.
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse BSR matrix.
bsrSortedValA – [in] array of nnzb blocks of the sparse BSR matrix.
bsrSortedRowPtrA – [in] array of mb+1 elements that point to the start of every block row of the sparse BSR matrix.
bsrSortedColIndA – [in] array of nnz containing the block column indices of the sparse BSR matrix.
blockDim – [in] block dimension of the sparse BSR matrix.
info – [in] structure that holds the information collected during the analysis step.
f – [in] array of m elements, holding the right-hand side.
x – [out] array of m elements, holding the solution.
policy – [in] HIPSPARSE_SOLVE_POLICY_NO_LEVEL or HIPSPARSE_SOLVE_POLICY_USE_LEVEL.
pBuffer – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, mb, nnzb, blockDim, descrA, alpha, bsrSortedValA, bsrSortedRowPtrA, bsrSortedColIndA, f or x is invalid.
HIPSPARSE_STATUS_ARCH_MISMATCH – the device is not supported.
HIPSPARSE_STATUS_INTERNAL_ERROR – an internal error occurred.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA == HIPSPARSE_OPERATION_CONJUGATE_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // A = ( 1.0  0.0  0.0  0.0 )
    //     ( 2.0  3.0  0.0  0.0 )
    //     ( 4.0  5.0  6.0  0.0 )
    //     ( 7.0  0.0  8.0  9.0 )
    //
    // with bsr_dim = 2
    //
    //      -------------------
    //   = | 1.0 0.0 | 0.0 0.0 |
    //     | 2.0 3.0 | 0.0 0.0 |
    //      -------------------
    //     | 4.0 5.0 | 6.0 0.0 |
    //     | 7.0 0.0 | 8.0 9.0 |
    //      -------------------

    // Number of rows (matrix must be square)
    const int m = 4;

    // Number of block rows and block columns
    const int mb = 2;
    const int nb = 2;

    // BSR block dimension
    const int bsr_dim = 2;

    // Number of non-zero blocks
    const int nnzb = 3;

    // BSR row pointers
    std::vector<int> hbsrRowPtr = {0, 1, 3};

    // BSR column indices
    std::vector<int> hbsrColInd = {0, 0, 1};

    // BSR values
    std::vector<double> hbsrVal = {1.0, 2.0, 0.0, 3.0, 4.0, 7.0, 5.0, 0.0, 6.0, 8.0, 0.0, 9.0};

    // Storage scheme of the BSR blocks
    hipsparseDirection_t dir = HIPSPARSE_DIRECTION_COLUMN;

    // Transposition of the matrix and rhs matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Solve policy
    hipsparseSolvePolicy_t solve_policy = HIPSPARSE_SOLVE_POLICY_USE_LEVEL;

    // Scalar alpha and beta
    double alpha = 3.7;

    std::vector<double> hx = {1, 2, 3, 4};
    std::vector<double> hy(m);

    // Offload data to device
    int*    dbsrRowPtr;
    int*    dbsrColInd;
    double* dbsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1)));
    HIP_CHECK(hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb));
    HIP_CHECK(hipMalloc((void**)&dbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * m));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(
        hipMemcpy(dbsrRowPtr, hbsrRowPtr.data(), sizeof(int) * (mb + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dbsrColInd, hbsrColInd.data(), sizeof(int) * nnzb, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(
        dbsrVal, hbsrVal.data(), sizeof(double) * nnzb * bsr_dim * bsr_dim, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx.data(), sizeof(double) * m, hipMemcpyHostToDevice));

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Matrix fill mode
    HIPSPARSE_CHECK(hipsparseSetMatFillMode(descr, HIPSPARSE_FILL_MODE_LOWER));

    // Matrix diagonal type
    HIPSPARSE_CHECK(hipsparseSetMatDiagType(descr, HIPSPARSE_DIAG_TYPE_UNIT));

    // Matrix info structure
    bsrsv2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateBsrsv2Info(&info));

    // Obtain required buffer size
    int buffer_size;
    HIPSPARSE_CHECK(hipsparseDbsrsv2_bufferSize(handle,
                                                dir,
                                                trans,
                                                mb,
                                                nnzb,
                                                descr,
                                                dbsrVal,
                                                dbsrRowPtr,
                                                dbsrColInd,
                                                bsr_dim,
                                                info,
                                                &buffer_size));

    // Allocate temporary buffer
    void* dbuffer;
    HIP_CHECK(hipMalloc(&dbuffer, buffer_size));

    // Perform analysis step
    HIPSPARSE_CHECK(hipsparseDbsrsv2_analysis(handle,
                                              dir,
                                              trans,
                                              mb,
                                              nnzb,
                                              descr,
                                              dbsrVal,
                                              dbsrRowPtr,
                                              dbsrColInd,
                                              bsr_dim,
                                              info,
                                              solve_policy,
                                              dbuffer));

    // Call dbsrsm to perform lower triangular solve LX = B
    HIPSPARSE_CHECK(hipsparseDbsrsv2_solve(handle,
                                           dir,
                                           trans,
                                           mb,
                                           nnzb,
                                           &alpha,
                                           descr,
                                           dbsrVal,
                                           dbsrRowPtr,
                                           dbsrColInd,
                                           bsr_dim,
                                           info,
                                           dx,
                                           dy,
                                           solve_policy,
                                           dbuffer));

    // Check for zero pivots
    int               pivot;
    hipsparseStatus_t status = hipsparseXbsrsv2_zeroPivot(handle, info, &pivot);

    if(status == HIPSPARSE_STATUS_ZERO_PIVOT)
    {
        std::cout << "Found zero pivot in matrix row " << pivot << std::endl;
    }

    // Copy results back to the host
    HIP_CHECK(hipMemcpy(hy.data(), dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    std::cout << "hy" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hy[i] << " ";
    }
    std::cout << std::endl;

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyBsrsv2Info(info));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dbsrRowPtr));
    HIP_CHECK(hipFree(dbsrColInd));
    HIP_CHECK(hipFree(dbsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

int main(int argc, char* argv[])
{
    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // A = ( 1.0  0.0  0.0  0.0 )
    //     ( 2.0  3.0  0.0  0.0 )
    //     ( 4.0  5.0  6.0  0.0 )
    //     ( 7.0  0.0  8.0  9.0 )
    //
    // with bsr_dim = 2
    //
    //      -------------------
    //   = | 1.0 0.0 | 0.0 0.0 |
    //     | 2.0 3.0 | 0.0 0.0 |
    //      -------------------
    //     | 4.0 5.0 | 6.0 0.0 |
    //     | 7.0 0.0 | 8.0 9.0 |
    //      -------------------

    // Number of rows (matrix must be square)
    const int m = 4;

    // Number of block rows and block columns
    const int mb = 2;
    const int nb = 2;

    // BSR block dimension
    const int bsr_dim = 2;

    // Number of non-zero blocks
    const int nnzb = 3;

    // BSR row pointers
    int hbsrRowPtr[] = {0, 1, 3};

    // BSR column indices
    int hbsrColInd[] = {0, 0, 1};

    // BSR values
    double hbsrVal[] = {1.0, 2.0, 0.0, 3.0, 4.0, 7.0, 5.0, 0.0, 6.0, 8.0, 0.0, 9.0};

    // Storage scheme of the BSR blocks
    hipsparseDirection_t dir = HIPSPARSE_DIRECTION_COLUMN;

    // Transposition of the matrix and rhs matrix
    hipsparseOperation_t trans = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // Solve policy
    hipsparseSolvePolicy_t solve_policy = HIPSPARSE_SOLVE_POLICY_USE_LEVEL;

    // Scalar alpha and beta
    double alpha = 3.7;

    double hx[] = {1, 2, 3, 4};
    double hy[m];

    // Offload data to device
    int*    dbsrRowPtr;
    int*    dbsrColInd;
    double* dbsrVal;
    double* dx;
    double* dy;

    HIP_CHECK(hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1)));
    HIP_CHECK(hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb));
    HIP_CHECK(hipMalloc((void**)&dbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(double) * m));
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(double) * m));

    HIP_CHECK(hipMemcpy(dbsrRowPtr, hbsrRowPtr, sizeof(int) * (mb + 1), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dbsrColInd, hbsrColInd, sizeof(int) * nnzb, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(
        dbsrVal, hbsrVal, sizeof(double) * nnzb * bsr_dim * bsr_dim, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx, sizeof(double) * m, hipMemcpyHostToDevice));

    // Matrix descriptor
    hipsparseMatDescr_t descr;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descr));

    // Matrix fill mode
    HIPSPARSE_CHECK(hipsparseSetMatFillMode(descr, HIPSPARSE_FILL_MODE_LOWER));

    // Matrix diagonal type
    HIPSPARSE_CHECK(hipsparseSetMatDiagType(descr, HIPSPARSE_DIAG_TYPE_UNIT));

    // Matrix info structure
    bsrsv2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateBsrsv2Info(&info));

    // Obtain required buffer size
    int buffer_size;
    HIPSPARSE_CHECK(hipsparseDbsrsv2_bufferSize(handle,
                                                dir,
                                                trans,
                                                mb,
                                                nnzb,
                                                descr,
                                                dbsrVal,
                                                dbsrRowPtr,
                                                dbsrColInd,
                                                bsr_dim,
                                                info,
                                                &buffer_size));

    // Allocate temporary buffer
    void* dbuffer;
    HIP_CHECK(hipMalloc(&dbuffer, buffer_size));

    // Perform analysis step
    HIPSPARSE_CHECK(hipsparseDbsrsv2_analysis(handle,
                                              dir,
                                              trans,
                                              mb,
                                              nnzb,
                                              descr,
                                              dbsrVal,
                                              dbsrRowPtr,
                                              dbsrColInd,
                                              bsr_dim,
                                              info,
                                              solve_policy,
                                              dbuffer));

    // Call dbsrsm to perform lower triangular solve LX = B
    HIPSPARSE_CHECK(hipsparseDbsrsv2_solve(handle,
                                           dir,
                                           trans,
                                           mb,
                                           nnzb,
                                           &alpha,
                                           descr,
                                           dbsrVal,
                                           dbsrRowPtr,
                                           dbsrColInd,
                                           bsr_dim,
                                           info,
                                           dx,
                                           dy,
                                           solve_policy,
                                           dbuffer));

    // Check for zero pivots
    int               pivot;
    hipsparseStatus_t status = hipsparseXbsrsv2_zeroPivot(handle, info, &pivot);

    if(status == HIPSPARSE_STATUS_ZERO_PIVOT)
    {
        printf("Found zero pivot in matrix row %d\n", pivot);
    }

    // Copy results back to the host
    HIP_CHECK(hipMemcpy(hy, dy, sizeof(double) * m, hipMemcpyDeviceToHost));

    printf("hy\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hy[i]);
    }
    printf("\n");

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroyBsrsv2Info(info));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dbsrRowPtr));
    HIP_CHECK(hipFree(dbsrColInd));
    HIP_CHECK(hipFree(dbsrVal));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

hipsparseXgemvi_bufferSize()#

hipsparseStatus_t hipsparseSgemvi_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDgemvi_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCgemvi_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, int *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZgemvi_bufferSize(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, int nnz, int *pBufferSizeInBytes)#

hipsparseXgemvi_bufferSize returns the size of the temporary storage buffer in bytes required by hipsparseXgemvi(). The temporary storage buffer must be allocated by the user.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the dense matrix.
n – [in] number of columns of the dense matrix.
nnz – [in] number of non-zero entries in the sparse vector.
pBufferSizeInBytes – [out] temporary storage buffer size.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, nnz or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA != HIPSPARSE_OPERATION_NON_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXgemvi()#

hipsparseStatus_t hipsparseSgemvi(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, const float *alpha, const float *A, int lda, int nnz, const float *x, const int *xInd, const float *beta, float *y, hipsparseIndexBase_t idxBase, void *pBuffer)#

hipsparseStatus_t hipsparseDgemvi(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, const double *alpha, const double *A, int lda, int nnz, const double *x, const int *xInd, const double *beta, double *y, hipsparseIndexBase_t idxBase, void *pBuffer)#

hipsparseStatus_t hipsparseCgemvi(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, const hipComplex *alpha, const hipComplex *A, int lda, int nnz, const hipComplex *x, const int *xInd, const hipComplex *beta, hipComplex *y, hipsparseIndexBase_t idxBase, void *pBuffer)#

hipsparseStatus_t hipsparseZgemvi(hipsparseHandle_t handle, hipsparseOperation_t transA, int m, int n, const hipDoubleComplex *alpha, const hipDoubleComplex *A, int lda, int nnz, const hipDoubleComplex *x, const int *xInd, const hipDoubleComplex *beta, hipDoubleComplex *y, hipsparseIndexBase_t idxBase, void *pBuffer)#

Dense matrix sparse vector multiplication.

hipsparseXgemvi multiplies the scalar \(\alpha\) with a dense \(m \times n\) matrix \(A\) and the sparse vector \(x\) and adds the result to the dense vector \(y\) that is multiplied by the scalar \(\beta\), such that

\[ y := \alpha \cdot op(A) \cdot x + \beta \cdot y, \]

with

\[ op(A) = \left\{ \begin{array}{ll} A, & \text{if transA == HIPSPARSE_OPERATION_NON_TRANSPOSE} \end{array} \right. \]

Performing the above operation involves two steps. First, the user calls hipsparseXgemvi_bufferSize() in order to determine the size of the temporary storage buffer. Next, the user allocates this temporary buffer and passes it to hipsparseXgemvi to complete the computation. Once all calls to hipsparseXgemvi are complete the temporary storage buffer can be freed.

Note

This function is non blocking and executed asynchronously with respect to the host. It may return before the actual computation has finished.

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix operation type.
m – [in] number of rows of the dense matrix.
n – [in] number of columns of the dense matrix.
alpha – [in] scalar \(\alpha\).
A – [in] pointer to the dense matrix.
lda – [in] leading dimension of the dense matrix
nnz – [in] number of non-zero entries in the sparse vector
x – [in] array of nnz elements containing the values of the sparse vector
xInd – [in] array of nnz elements containing the indices of the sparse vector
beta – [in] scalar \(\beta\).
y – [inout] array of m elements ( \(op(A) == A\)) or n elements ( \(op(A) == A^T\) or \(op(A) == A^H\)).
idxBase – [in] HIPSPARSE_INDEX_BASE_ZERO or HIPSPARSE_INDEX_BASE_ONE.
pBuffer – [in] temporary storage buffer

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, lda, nnz, alpha, A, x, xInd, beta, y or pBuffer is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA != HIPSPARSE_OPERATION_NON_TRANSPOSE or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    hipsparseOperation_t opA     = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseIndexBase_t idxBase = HIPSPARSE_INDEX_BASE_ZERO;

    // Scalar alpha and beta
    float alpha = 1.0f;
    float beta  = 1.0f;

    const int m   = 4; // Number of rows of A
    const int n   = 4; // Number of columns of A
    const int lda = m; // leading dimension of A

    // A = 1 2 3 4
    //     5 6 7 8
    //     2 4 6 8
    //     4 3 2 1
    std::vector<float> hA = {1.0f,
                             5.0f,
                             2.0f,
                             4.0f,
                             2.0f,
                             6.0f,
                             4.0f,
                             3.0f,
                             3.0f,
                             7.0f,
                             6.0f,
                             2.0f,
                             4.0f,
                             8.0f,
                             8.0f,
                             1.0f};

    // Sparse vector x
    int                nnz   = 2;
    std::vector<int>   hxInd = {0, 2};
    std::vector<float> hx    = {10.0f, 11.0f};

    // Dense vector y
    std::vector<float> hy = {1.0f, 2.0f, 3.0f, 4.0f};

    // Device data
    float* dA = nullptr;
    HIP_CHECK(hipMalloc((void**)&dA, sizeof(float) * m * n));

    int*   dxInd = nullptr;
    float* dx    = nullptr;
    HIP_CHECK(hipMalloc((void**)&dxInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(float) * nnz));

    float* dy = nullptr;
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(float) * m));

    // Copy data from host to device
    HIP_CHECK(hipMemcpy(dA, hA.data(), sizeof(float) * m * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dxInd, hxInd.data(), sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx.data(), sizeof(float) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy.data(), sizeof(float) * m, hipMemcpyHostToDevice));

    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // Call hipsparseSgemvi to perform y = alpha * A * x + beta * y
    int bufferSize = 0;
    HIPSPARSE_CHECK(hipsparseSgemvi_bufferSize(handle, opA, m, n, nnz, &bufferSize));

    void* dbuffer = nullptr;
    HIP_CHECK(hipMalloc((void**)&dbuffer, bufferSize));

    HIPSPARSE_CHECK(hipsparseSgemvi(
        handle, opA, m, n, &alpha, dA, lda, nnz, dx, dxInd, &beta, dy, idxBase, dbuffer));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy.data(), dy, sizeof(float) * m, hipMemcpyDeviceToHost));

    // Print the result (optional)
    std::cout << "hy" << std::endl;
    for(int i = 0; i < m; i++)
    {
        std::cout << hy[i] << " ";
    }
    std::cout << "" << std::endl;

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dA));
    HIP_CHECK(hipFree(dxInd));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

int main(int argc, char* argv[])
{
    hipsparseOperation_t opA     = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseIndexBase_t idxBase = HIPSPARSE_INDEX_BASE_ZERO;

    // Scalar alpha and beta
    float alpha = 1.0f;
    float beta  = 1.0f;

    const int m   = 4; // Number of rows of A
    const int n   = 4; // Number of columns of A
    const int lda = m; // leading dimension of A

    // A = 1 2 3 4
    //     5 6 7 8
    //     2 4 6 8
    //     4 3 2 1
    float hA[16] = {1.0f,
                    5.0f,
                    2.0f,
                    4.0f,
                    2.0f,
                    6.0f,
                    4.0f,
                    3.0f,
                    3.0f,
                    7.0f,
                    6.0f,
                    2.0f,
                    4.0f,
                    8.0f,
                    8.0f,
                    1.0f};

    // Sparse vector x
    int   nnz      = 2;
    int   hxInd[2] = {0, 2};
    float hx[2]    = {10.0f, 11.0f};

    // Dense vector y
    float hy[4] = {1.0f, 2.0f, 3.0f, 4.0f};

    // Device data
    float* dA = NULL;
    HIP_CHECK(hipMalloc((void**)&dA, sizeof(float) * m * n));

    int*   dxInd = NULL;
    float* dx    = NULL;
    HIP_CHECK(hipMalloc((void**)&dxInd, sizeof(int) * nnz));
    HIP_CHECK(hipMalloc((void**)&dx, sizeof(float) * nnz));

    float* dy = NULL;
    HIP_CHECK(hipMalloc((void**)&dy, sizeof(float) * m));

    // Copy data from host to device
    HIP_CHECK(hipMemcpy(dA, hA, sizeof(float) * m * n, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dxInd, hxInd, sizeof(int) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dx, hx, sizeof(float) * nnz, hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dy, hy, sizeof(float) * m, hipMemcpyHostToDevice));

    // hipSPARSE handle
    hipsparseHandle_t handle;
    HIPSPARSE_CHECK(hipsparseCreate(&handle));

    // Call hipsparseSgemvi to perform y = alpha * A * x + beta * y
    int bufferSize = 0;
    HIPSPARSE_CHECK(hipsparseSgemvi_bufferSize(handle, opA, m, n, nnz, &bufferSize));

    void* dbuffer = NULL;
    HIP_CHECK(hipMalloc((void**)&dbuffer, bufferSize));

    HIPSPARSE_CHECK(hipsparseSgemvi(
        handle, opA, m, n, &alpha, dA, lda, nnz, dx, dxInd, &beta, dy, idxBase, dbuffer));

    // Copy result back to host
    HIP_CHECK(hipMemcpy(hy, dy, sizeof(float) * m, hipMemcpyDeviceToHost));

    // Print the result (optional)
    printf("hy\n");
    for(int i = 0; i < m; i++)
    {
        printf("%f ", hy[i]);
    }
    printf("\n");

    // Clear hipSPARSE
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    // Clear device memory
    HIP_CHECK(hipFree(dA));
    HIP_CHECK(hipFree(dxInd));
    HIP_CHECK(hipFree(dx));
    HIP_CHECK(hipFree(dy));
    HIP_CHECK(hipFree(dbuffer));

    return 0;
}

Sparse level 2 functions

Contents

Sparse level 2 functions#

hipsparseXcsrmv()#

hipsparseXcsrsv2_zeroPivot()#

hipsparseXcsrsv2_bufferSize()#

hipsparseXcsrsv2_bufferSizeExt()#

hipsparseXcsrsv2_analysis()#

hipsparseXcsrsv2_solve()#

hipsparseXhybmv()#

hipsparseXbsrmv()#

hipsparseXbsrxmv()#

hipsparseXbsrsv2_zeroPivot()#

hipsparseXbsrsv2_bufferSize()#

hipsparseXbsrsv2_bufferSizeExt()#

hipsparseXbsrsv2_analysis()#

hipsparseXbsrsv2_solve()#

hipsparseXgemvi_bufferSize()#

hipsparseXgemvi()#