Sparse extra functions

Sparse extra functions#

This module contains all sparse extra routines.

The sparse extra routines describe operations that manipulate sparse matrices.

hipsparseXcsrgeamNnz()#

hipsparseStatus_t hipsparseXcsrgeamNnz(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr)#

hipsparseXcsrgeamNnz computes the total CSR non-zero elements and the CSR row offsets, that point to the start of every row of the sparse CSR matrix, of the resulting matrix \(C\). It is assumed that csrRowPtrC has been allocated with size m+1. The desired index base in the output CSR matrix is set in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase().

For full code example, see hipsparseScsrgeam().

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

As indicated, nnzTotalDevHostPtr can point either to host or device memory. This is controlled by setting the pointer mode. See hipsparseSetPointerMode().

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 HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(A\), \(B\) and \(C\). Must be non-negative.
n – [in] number of columns of the sparse CSR matrix \(A\), \(B\) and \(C\). Must be non-negative.
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\). Must be non-negative.
csrRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(A\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\). Must be non-negative.
csrRowPtrB – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(B\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrRowPtrC – [out] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
nnzTotalDevHostPtr – [out] pointer to the number of non-zero entries of the sparse CSR matrix \(C\). nnzTotalDevHostPtr can be a host or device pointer.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descrA, descrB or descrC is nullptr, m, n, nnzA or nnzB is negative, or csrRowPtrA, csrColIndA, csrRowPtrB, csrColIndB, csrRowPtrC or nnzTotalDevHostPtr is nullptr.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgeam()#

hipsparseStatus_t hipsparseScsrgeam(hipsparseHandle_t handle, int m, int n, const float *alpha, const hipsparseMatDescr_t descrA, int nnzA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const float *beta, const hipsparseMatDescr_t descrB, int nnzB, const float *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, float *csrValC, int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseDcsrgeam(hipsparseHandle_t handle, int m, int n, const double *alpha, const hipsparseMatDescr_t descrA, int nnzA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const double *beta, const hipsparseMatDescr_t descrB, int nnzB, const double *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, double *csrValC, int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseCcsrgeam(hipsparseHandle_t handle, int m, int n, const hipComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, hipComplex *csrValC, int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseZcsrgeam(hipsparseHandle_t handle, int m, int n, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipDoubleComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipDoubleComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipDoubleComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrValC, int *csrRowPtrC, int *csrColIndC)#

Sparse matrix sparse matrix addition using CSR storage format.

hipsparseXcsrgeam multiplies the scalar \(\alpha\) with the sparse \(m \times n\) matrix \(A\), defined in CSR storage format, multiplies the scalar \(\beta\) with the sparse \(m \times n\) matrix \(B\), defined in CSR storage format, and adds both resulting matrices to obtain the sparse \(m \times n\) matrix \(C\), defined in CSR storage format, such that

\[ C := \alpha \cdot A + \beta \cdot B. \]

This computation involves a multi step process. First the user must allocate csrRowPtrC to have size m+1. The user then calls hipsparseXcsrgeamNnz which fills in the csrRowPtrC array as well as computes the total number of nonzeros in \(C\), nnzC. The user then allocates both arrays csrColIndC and csrValC to have size nnzC and calls hipsparseXcsrgeam to complete the computation. The desired index base in the output CSR matrix \(C\) is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

Both scalars \(\alpha\) and \(beta\) have to be valid.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

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.
m – [in] number of rows of the sparse CSR matrix \(A\), \(B\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(A\), \(B\) and \(C\).
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrValA – [in] array of nnzA elements of the sparse CSR matrix \(A\).
csrRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(A\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
beta – [in] scalar \(\beta\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrValB – [in] array of nnzB elements of the sparse CSR matrix \(B\).
csrRowPtrB – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(B\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrValC – [out] array of elements of the sparse CSR matrix \(C\).
csrRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
csrColIndC – [out] array of elements containing the column indices of the sparse CSR matrix \(C\).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, nnzA, nnzB, alpha, descrA, csrValA, csrRowPtrA, csrColIndA, beta, descrB, csrValB, csrRowPtrB, csrColIndB, descrC, csrValC, csrRowPtrC or csrColIndC is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int n    = 4;
    const int nnzA = 9;
    const int nnzB = 6;

    float alpha{1.0f};
    float beta{1.0f};

    // A, B, and C are m×n

    // A
    // 1 0 0 2
    // 3 4 0 0
    // 5 6 7 8
    // 0 0 9 0
    std::vector<int>   hcsrRowPtrA = {0, 2, 4, 8, 9};
    std::vector<int>   hcsrColIndA = {0, 3, 0, 1, 0, 1, 2, 3, 2};
    std::vector<float> hcsrValA    = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f, 9.0f};

    // B
    // 0 1 0 0
    // 1 0 1 0
    // 0 1 0 1
    // 0 0 1 0
    std::vector<int>   hcsrRowPtrB = {0, 1, 3, 5, 6};
    std::vector<int>   hcsrColIndB = {1, 0, 2, 1, 3, 2};
    std::vector<float> hcsrValB    = {1.0f, 1.0f, 1.0f, 1.0f, 1.0f, 1.0f};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtrA, hcsrRowPtrA.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndA, hcsrColIndA.data(), nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA.data(), nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrRowPtrB, hcsrRowPtrB.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndB, hcsrColIndB.data(), nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB.data(), nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgeamNnz(handle,
                                         m,
                                         n,
                                         descrA,
                                         nnzA,
                                         dcsrRowPtrA,
                                         dcsrColIndA,
                                         descrB,
                                         nnzB,
                                         dcsrRowPtrB,
                                         dcsrColIndB,
                                         descrC,
                                         dcsrRowPtrC,
                                         &nnzC));

    int*   dcsrColIndC = nullptr;
    float* dcsrValC    = nullptr;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgeam(handle,
                                      m,
                                      n,
                                      &alpha,
                                      descrA,
                                      nnzA,
                                      dcsrValA,
                                      dcsrRowPtrA,
                                      dcsrColIndA,
                                      &beta,
                                      descrB,
                                      nnzB,
                                      dcsrValB,
                                      dcsrRowPtrB,
                                      dcsrColIndB,
                                      descrC,
                                      dcsrValC,
                                      dcsrRowPtrC,
                                      dcsrColIndC));

    std::vector<int>   hcsrRowPtrC(m + 1);
    std::vector<int>   hcsrColIndC(nnzC);
    std::vector<float> hcsrValC(nnzC);

    // Copy back to the host
    HIP_CHECK(
        hipMemcpy(hcsrRowPtrC.data(), dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(
        hipMemcpy(hcsrColIndC.data(), dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC.data(), dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    std::cout << "C" << std::endl;
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        std::vector<float> temp(n, 0.0f);
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

        for(int j = 0; j < n; j++)
        {
            std::cout << temp[j] << " ";
        }
        std::cout << std::endl;
    }
    std::cout << std::endl;

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int n    = 4;
    const int nnzA = 9;
    const int nnzB = 6;

    float alpha = 1.0;
    float beta  = 1.0;

    // A, B, and C are m×n

    // A
    // 1 0 0 2
    // 3 4 0 0
    // 5 6 7 8
    // 0 0 9 0
    int   hcsrRowPtrA[] = {0, 2, 4, 8, 9};
    int   hcsrColIndA[] = {0, 3, 0, 1, 0, 1, 2, 3, 2};
    float hcsrValA[]    = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0};

    // B
    // 0 1 0 0
    // 1 0 1 0
    // 0 1 0 1
    // 0 0 1 0
    int   hcsrRowPtrB[] = {0, 1, 3, 5, 6};
    int   hcsrColIndB[] = {1, 0, 2, 1, 3, 2};
    float hcsrValB[]    = {1.0, 1.0, 1.0, 1.0, 1.0, 1.0};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(hipMemcpy(dcsrRowPtrA, hcsrRowPtrA, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndA, hcsrColIndA, nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA, nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrRowPtrB, hcsrRowPtrB, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndB, hcsrColIndB, nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB, nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgeamNnz(handle,
                                         m,
                                         n,
                                         descrA,
                                         nnzA,
                                         dcsrRowPtrA,
                                         dcsrColIndA,
                                         descrB,
                                         nnzB,
                                         dcsrRowPtrB,
                                         dcsrColIndB,
                                         descrC,
                                         dcsrRowPtrC,
                                         &nnzC));

    int*   dcsrColIndC = NULL;
    float* dcsrValC    = NULL;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgeam(handle,
                                      m,
                                      n,
                                      &alpha,
                                      descrA,
                                      nnzA,
                                      dcsrValA,
                                      dcsrRowPtrA,
                                      dcsrColIndA,
                                      &beta,
                                      descrB,
                                      nnzB,
                                      dcsrValB,
                                      dcsrRowPtrB,
                                      dcsrColIndB,
                                      descrC,
                                      dcsrValC,
                                      dcsrRowPtrC,
                                      dcsrColIndC));

    int*  hcsrRowPtrC = (int*)malloc((m + 1) * sizeof(int));
    int*  hcsrColIndC = (int*)malloc((nnzC) * sizeof(int));
    float hcsrValC[nnzC];

    // Copy back to the host
    HIP_CHECK(hipMemcpy(hcsrRowPtrC, dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrColIndC, dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC, dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    printf("C\n");
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        float* temp = (float*)malloc(n * sizeof(float));
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

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

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

program example_hipsparse_csrgeam
    use iso_c_binding
    implicit none

    ! HIP
    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer(c_int) :: 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(c_int) :: 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(c_int) :: 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, parameter :: hipMemcpyHostToDevice = 1
    integer, parameter :: hipMemcpyDeviceToHost = 2

    ! hipSPARSE
    interface
        function hipsparseCreate(handle) &
                bind(c, name = 'hipsparseCreate')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreate
            type(c_ptr) :: handle
        end function hipsparseCreate

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

        function hipsparseCreateMatDescr(descr) &
                bind(c, name = 'hipsparseCreateMatDescr')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreateMatDescr
            type(c_ptr) :: descr
        end function hipsparseCreateMatDescr

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

        function hipsparseXcsrgeamNnz(handle, m, n, descrA, nnzA, csrRowPtrA, csrColIndA, descrB, &
                                      nnzB, csrRowPtrB, csrColIndB, descrC, csrRowPtrC, nnzTotalDevHostPtr) &
                bind(c, name = 'hipsparseXcsrgeamNnz')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseXcsrgeamNnz
            type(c_ptr), value :: handle
            integer(c_int), value :: m
            integer(c_int), value :: n
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrRowPtrA
            type(c_ptr), intent(in), value :: csrColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrRowPtrB
            type(c_ptr), intent(in), value :: csrColIndB
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrRowPtrC
            type(c_ptr), value :: nnzTotalDevHostPtr
        end function hipsparseXcsrgeamNnz

        function hipsparseScsrgeam(handle, m, n, alpha, descrA, nnzA, csrSortedValA, csrSortedRowPtrA, &
                                   csrSortedColIndA, beta, descrB, nnzB, csrSortedValB, csrSortedRowPtrB, &
                                   csrSortedColIndB, descrC, csrSortedValC, csrSortedRowPtrC, csrSortedColIndC) &
                bind(c, name = 'hipsparseScsrgeam')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseScsrgeam
            type(c_ptr), value :: handle
            integer(c_int), value :: m
            integer(c_int), value :: n
            type(c_ptr), intent(in), value :: alpha
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrSortedValA
            type(c_ptr), intent(in), value :: csrSortedRowPtrA
            type(c_ptr), intent(in), value :: csrSortedColIndA
            type(c_ptr), intent(in), value :: beta
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrSortedValB
            type(c_ptr), intent(in), value :: csrSortedRowPtrB
            type(c_ptr), intent(in), value :: csrSortedColIndB
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrSortedValC
            type(c_ptr), value :: csrSortedRowPtrC
            type(c_ptr), value :: csrSortedColIndC
        end function hipsparseScsrgeam
    end interface

    ! Variables
    type(c_ptr) :: handle
    type(c_ptr) :: descrA, descrB, descrC
    integer :: i, stat
    
    ! Matrix addition: C = alpha * A + beta * B (no transpose!)
    integer, parameter :: m = 4
    integer, parameter :: n = 4
    integer, parameter :: nnzA = 9
    integer, parameter :: nnzB = 6
    integer(c_int), target :: nnzC
    
    ! Matrix A (4x4)
    integer, dimension(m+1), target :: hcsrRowPtrA = (/0, 2, 4, 8, 9/)
    integer, dimension(nnzA), target :: hcsrColIndA = (/0, 3, 0, 1, 0, 1, 2, 3, 2/)
    real(c_float), dimension(nnzA), target :: hcsrValA = (/1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0/)
    
    ! Matrix B (4x4)
    integer, dimension(m+1), target :: hcsrRowPtrB = (/0, 1, 3, 5, 6/)
    integer, dimension(nnzB), target :: hcsrColIndB = (/1, 0, 2, 1, 3, 2/)
    real(c_float), dimension(nnzB), target :: hcsrValB = (/1.0, 1.0, 1.0, 1.0, 1.0, 1.0/)
    
    ! Matrix C (will be allocated after nnzC is determined)
    integer, dimension(:), allocatable, target :: hcsrRowPtrC
    integer, dimension(:), allocatable, target :: hcsrColIndC
    real(c_float), dimension(:), allocatable, target :: hcsrValC
    
    ! Scalar values
    real(c_float), target :: alpha = 1.0
    real(c_float), target :: beta = 1.0
    
    ! Device pointers
    type(c_ptr) :: dcsrRowPtrA, dcsrColIndA, dcsrValA
    type(c_ptr) :: dcsrRowPtrB, dcsrColIndB, dcsrValB
    type(c_ptr) :: dcsrRowPtrC, dcsrColIndC, dcsrValC

    ! Create hipSPARSE handle
    stat = hipsparseCreate(handle)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseCreate failed'
        stop
    end if

    ! Create matrix descriptors
    stat = hipsparseCreateMatDescr(descrA)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseCreateMatDescr descrA failed'
        stop
    end if

    stat = hipsparseCreateMatDescr(descrB)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseCreateMatDescr descrB failed'
        stop
    end if

    stat = hipsparseCreateMatDescr(descrC)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseCreateMatDescr descrC failed'
        stop
    end if

    ! Allocate device memory for A and B
    stat = hipMalloc(dcsrRowPtrA, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrB, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrC, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop

    ! Copy A and B to device
    stat = hipMemcpy(dcsrRowPtrA, c_loc(hcsrRowPtrA), int((m + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndA, c_loc(hcsrColIndA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValA, c_loc(hcsrValA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    stat = hipMemcpy(dcsrRowPtrB, c_loc(hcsrRowPtrB), int((m + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndB, c_loc(hcsrColIndB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValB, c_loc(hcsrValB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    ! Determine nnzC
    stat = hipsparseXcsrgeamNnz(handle, &
                                m, &
                                n, &
                                descrA, &
                                nnzA, &
                                dcsrRowPtrA, &
                                dcsrColIndA, &
                                descrB, &
                                nnzB, &
                                dcsrRowPtrB, &
                                dcsrColIndB, &
                                descrC, &
                                dcsrRowPtrC, &
                                c_loc(nnzC))
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseXcsrgeamNnz failed'
        stop
    end if

    ! Allocate device memory for C
    stat = hipMalloc(dcsrColIndC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop

    ! Perform matrix addition: C = alpha * A + beta * B
    stat = hipsparseScsrgeam(handle, &
                             m, &
                             n, &
                             c_loc(alpha), &
                             descrA, &
                             nnzA, &
                             dcsrValA, &
                             dcsrRowPtrA, &
                             dcsrColIndA, &
                             c_loc(beta), &
                             descrB, &
                             nnzB, &
                             dcsrValB, &
                             dcsrRowPtrB, &
                             dcsrColIndB, &
                             descrC, &
                             dcsrValC, &
                             dcsrRowPtrC, &
                             dcsrColIndC)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseScsrgeam failed'
        stop
    end if

    ! Allocate host memory for C
    allocate(hcsrRowPtrC(m+1))
    allocate(hcsrColIndC(nnzC))
    allocate(hcsrValC(nnzC))

    ! Copy result back to host
    stat = hipMemcpy(c_loc(hcsrRowPtrC), dcsrRowPtrC, int((m + 1) * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrColIndC), dcsrColIndC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrValC), dcsrValC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop

    ! Print result
    write(*,*) 'Matrix C (result of A + B):'
    write(*,*) 'nnzC =', nnzC
    write(*,*) 'csrRowPtrC:'
    do i = 1, m + 1
        write(*,*) hcsrRowPtrC(i)
    end do
    write(*,*) 'csrColIndC:'
    do i = 1, nnzC
        write(*,*) hcsrColIndC(i)
    end do
    write(*,*) 'csrValC:'
    do i = 1, nnzC
        write(*,*) hcsrValC(i)
    end do

    ! Clean up
    deallocate(hcsrRowPtrC)
    deallocate(hcsrColIndC)
    deallocate(hcsrValC)

    stat = hipFree(dcsrRowPtrA)
    stat = hipFree(dcsrColIndA)
    stat = hipFree(dcsrValA)
    stat = hipFree(dcsrRowPtrB)
    stat = hipFree(dcsrColIndB)
    stat = hipFree(dcsrValB)
    stat = hipFree(dcsrRowPtrC)
    stat = hipFree(dcsrColIndC)
    stat = hipFree(dcsrValC)

    stat = hipsparseDestroyMatDescr(descrA)
    stat = hipsparseDestroyMatDescr(descrB)
    stat = hipsparseDestroyMatDescr(descrC)
    stat = hipsparseDestroy(handle)

end program example_hipsparse_csrgeam

hipsparseXcsrgeam2_bufferSizeExt()#

hipsparseStatus_t hipsparseScsrgeam2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const float *alpha, const hipsparseMatDescr_t descrA, int nnzA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const float *beta, const hipsparseMatDescr_t descrB, int nnzB, const float *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, const float *csrSortedValC, const int *csrSortedRowPtrC, const int *csrSortedColIndC, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDcsrgeam2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const double *alpha, const hipsparseMatDescr_t descrA, int nnzA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const double *beta, const hipsparseMatDescr_t descrB, int nnzB, const double *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, const double *csrSortedValC, const int *csrSortedRowPtrC, const int *csrSortedColIndC, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCcsrgeam2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const hipComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipComplex *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, const hipComplex *csrSortedValC, const int *csrSortedRowPtrC, const int *csrSortedColIndC, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZcsrgeam2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipDoubleComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipDoubleComplex *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, const hipDoubleComplex *csrSortedValC, const int *csrSortedRowPtrC, const int *csrSortedColIndC, size_t *pBufferSizeInBytes)#

hipsparseXcsrgeam2_bufferSizeExt returns the size of the temporary storage buffer in bytes that is required by hipsparseXcsrgeam2Nnz() and hipsparseXcsrgeam2(). The temporary storage buffer must be allocated by the user.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(A\), \(B\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(A\), \(B\) and \(C\).
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrSortedValA – [in] array of nnzA elements of the sparse CSR matrix \(A\).
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(A\).
csrSortedColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
beta – [in] scalar \(\beta\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrSortedValB – [in] array of nnzB elements of the sparse CSR matrix \(B\).
csrSortedRowPtrB – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(B\).
csrSortedColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrSortedValC – [out] array of elements of the sparse CSR matrix \(C\).
csrSortedRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
csrSortedColIndC – [out] array of elements containing the column indices of the sparse CSR matrix \(C\).
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXcsrgeam2Nnz() and hipsparseXcsrgeam2().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, nnzA, nnzB, alpha, descrA, csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, beta, descrB, csrSortedValB, csrSortedRowPtrB, csrSortedColIndB, descrC, csrSortedValC, csrSortedRowPtrC, csrSortedColIndC, or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgeam2Nnz()#

hipsparseStatus_t hipsparseXcsrgeam2Nnz(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, int nnzA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, int *csrSortedRowPtrC, int *nnzTotalDevHostPtr, void *workspace)#

hipsparseXcsrgeam2Nnz computes the total CSR non-zero elements and the CSR row offsets, that point to the start of every row of the sparse CSR matrix, of the resulting matrix \(C\). It is assumed that csrRowPtrC has been allocated with size m+1. The required buffer size can be obtained by hipsparseXcsrgeam2_bufferSizeExt(). The desired index base in the output CSR matrix \(C\) is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

As indicated, nnzTotalDevHostPtr can point either to host or device memory. This is controlled by setting the pointer mode. See hipsparseSetPointerMode().

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 HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(A\), \(B\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(A\), \(B\) and \(C\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(A\).
csrSortedColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrSortedRowPtrB – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(B\).
csrSortedColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrSortedRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
nnzTotalDevHostPtr – [out] pointer to the number of non-zero entries of the sparse CSR matrix \(C\). nnzTotalDevHostPtr can be a host or device pointer.
workspace – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, nnzA, nnzB, descrA, csrSortedRowPtrA, csrSortedColIndA, descrB, csrSortedRowPtrB, csrSortedColIndB, descrC, csrSortedRowPtrC or nnzTotalDevHostPtr is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgeam2()#

hipsparseStatus_t hipsparseScsrgeam2(hipsparseHandle_t handle, int m, int n, const float *alpha, const hipsparseMatDescr_t descrA, int nnzA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const float *beta, const hipsparseMatDescr_t descrB, int nnzB, const float *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, float *csrSortedValC, int *csrSortedRowPtrC, int *csrSortedColIndC, void *pBuffer)#

hipsparseStatus_t hipsparseDcsrgeam2(hipsparseHandle_t handle, int m, int n, const double *alpha, const hipsparseMatDescr_t descrA, int nnzA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const double *beta, const hipsparseMatDescr_t descrB, int nnzB, const double *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, double *csrSortedValC, int *csrSortedRowPtrC, int *csrSortedColIndC, void *pBuffer)#

hipsparseStatus_t hipsparseCcsrgeam2(hipsparseHandle_t handle, int m, int n, const hipComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipComplex *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, hipComplex *csrSortedValC, int *csrSortedRowPtrC, int *csrSortedColIndC, void *pBuffer)#

hipsparseStatus_t hipsparseZcsrgeam2(hipsparseHandle_t handle, int m, int n, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, const hipDoubleComplex *beta, const hipsparseMatDescr_t descrB, int nnzB, const hipDoubleComplex *csrSortedValB, const int *csrSortedRowPtrB, const int *csrSortedColIndB, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrSortedValC, int *csrSortedRowPtrC, int *csrSortedColIndC, void *pBuffer)#

Sparse matrix sparse matrix addition using CSR storage format.

hipsparseXcsrgeam2 multiplies the scalar \(\alpha\) with the sparse \(m \times n\) matrix \(A\), defined in CSR storage format, multiplies the scalar \(\beta\) with the sparse \(m \times n\) matrix \(B\), defined in CSR storage format, and adds both resulting matrices to obtain the sparse \(m \times n\) matrix \(C\), defined in CSR storage format, such that

\[ C := \alpha \cdot A + \beta \cdot B. \]

This computation involves a multi step process. First the user must call hipsparseXcsrgeam2_bufferSizeExt() in order to determine the required user allocated temporary buffer size. The user then allocates this buffer and also allocates csrRowPtrC to have size m+1. Both the temporary storage buffer and csrRowPtrC array are then passed to hipsparseXcsrgeam2Nnz which fills in the csrRowPtrC array as well as computes the total number of nonzeros in C, nnzC. The user then allocates both arrays csrColIndC and csrValC to have size nnzC and calls hipsparseXcsrgeam2 to complete the computation. The desired index base in the output CSR matrix C is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

Both scalars \(\alpha\) and \(beta\) have to be valid.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

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.
m – [in] number of rows of the sparse CSR matrix \(A\), \(B\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(A\), \(B\) and \(C\).
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrSortedValA – [in] array of nnzA elements of the sparse CSR matrix \(A\).
csrSortedRowPtrA – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(A\).
csrSortedColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
beta – [in] scalar \(\beta\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrSortedValB – [in] array of nnzB elements of the sparse CSR matrix \(B\).
csrSortedRowPtrB – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(B\).
csrSortedColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrSortedValC – [out] array of elements of the sparse CSR matrix \(C\).
csrSortedRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
csrSortedColIndC – [out] array of elements containing the column indices of the sparse CSR matrix \(C\).
pBuffer – [in] temporary storage buffer allocated by the user.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, nnzA, nnzB, alpha, descrA, csrSortedValA, csrSortedRowPtrA, csrSortedColIndA, beta, descrB, csrSortedValB, csrSortedRowPtrB, csrSortedColIndB, descrC, csrSortedValC, csrSortedRowPtrC, csrSortedColIndC or pBuffer is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int n    = 4;
    const int nnzA = 9;
    const int nnzB = 6;

    float alpha{1.0f};
    float beta{1.0f};

    // A, B, and C are m×n

    // A
    // 1 0 0 2
    // 3 4 0 0
    // 5 6 7 8
    // 0 0 9 0
    std::vector<int>   hcsrRowPtrA = {0, 2, 4, 8, 9};
    std::vector<int>   hcsrColIndA = {0, 3, 0, 1, 0, 1, 2, 3, 2};
    std::vector<float> hcsrValA    = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f, 9.0f};

    // B
    // 0 1 0 0
    // 1 0 1 0
    // 0 1 0 1
    // 0 0 1 0
    std::vector<int>   hcsrRowPtrB = {0, 1, 3, 5, 6};
    std::vector<int>   hcsrColIndB = {1, 0, 2, 1, 3, 2};
    std::vector<float> hcsrValB    = {1.0f, 1.0f, 1.0f, 1.0f, 1.0f, 1.0f};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtrA, hcsrRowPtrA.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndA, hcsrColIndA.data(), nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA.data(), nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrRowPtrB, hcsrRowPtrB.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndB, hcsrColIndB.data(), nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB.data(), nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    size_t bufferSize;
    HIPSPARSE_CHECK(hipsparseScsrgeam2_bufferSizeExt(handle,
                                                     m,
                                                     n,
                                                     &alpha,
                                                     descrA,
                                                     nnzA,
                                                     dcsrValA,
                                                     dcsrRowPtrA,
                                                     dcsrColIndA,
                                                     &beta,
                                                     descrB,
                                                     nnzB,
                                                     dcsrValB,
                                                     dcsrRowPtrB,
                                                     dcsrColIndB,
                                                     descrC,
                                                     nullptr,
                                                     dcsrRowPtrC,
                                                     nullptr,
                                                     &bufferSize));

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

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgeam2Nnz(handle,
                                          m,
                                          n,
                                          descrA,
                                          nnzA,
                                          dcsrRowPtrA,
                                          dcsrColIndA,
                                          descrB,
                                          nnzB,
                                          dcsrRowPtrB,
                                          dcsrColIndB,
                                          descrC,
                                          dcsrRowPtrC,
                                          &nnzC,
                                          dbuffer));

    int*   dcsrColIndC = nullptr;
    float* dcsrValC    = nullptr;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgeam2(handle,
                                       m,
                                       n,
                                       &alpha,
                                       descrA,
                                       nnzA,
                                       dcsrValA,
                                       dcsrRowPtrA,
                                       dcsrColIndA,
                                       &beta,
                                       descrB,
                                       nnzB,
                                       dcsrValB,
                                       dcsrRowPtrB,
                                       dcsrColIndB,
                                       descrC,
                                       dcsrValC,
                                       dcsrRowPtrC,
                                       dcsrColIndC,
                                       dbuffer));

    std::vector<int>   hcsrRowPtrC(m + 1);
    std::vector<int>   hcsrColIndC(nnzC);
    std::vector<float> hcsrValC(nnzC);

    // Copy back to the host
    HIP_CHECK(
        hipMemcpy(hcsrRowPtrC.data(), dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(
        hipMemcpy(hcsrColIndC.data(), dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC.data(), dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    std::cout << "C" << std::endl;
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        std::vector<float> temp(n, 0.0f);
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

        for(int j = 0; j < n; j++)
        {
            std::cout << temp[j] << " ";
        }
        std::cout << std::endl;
    }
    std::cout << std::endl;

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIP_CHECK(hipFree(dbuffer));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int n    = 4;
    const int nnzA = 9;
    const int nnzB = 6;

    float alpha = 1.0;
    float beta  = 1.0;

    // A, B, and C are m×n

    // A
    // 1 0 0 2
    // 3 4 0 0
    // 5 6 7 8
    // 0 0 9 0
    int   hcsrRowPtrA[] = {0, 2, 4, 8, 9};
    int   hcsrColIndA[] = {0, 3, 0, 1, 0, 1, 2, 3, 2};
    float hcsrValA[]    = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0};

    // B
    // 0 1 0 0
    // 1 0 1 0
    // 0 1 0 1
    // 0 0 1 0
    int   hcsrRowPtrB[] = {0, 1, 3, 5, 6};
    int   hcsrColIndB[] = {1, 0, 2, 1, 3, 2};
    float hcsrValB[]    = {1.0, 1.0, 1.0, 1.0, 1.0, 1.0};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(hipMemcpy(dcsrRowPtrA, hcsrRowPtrA, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndA, hcsrColIndA, nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA, nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrRowPtrB, hcsrRowPtrB, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndB, hcsrColIndB, nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB, nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    size_t bufferSize;
    HIPSPARSE_CHECK(hipsparseScsrgeam2_bufferSizeExt(handle,
                                                     m,
                                                     n,
                                                     &alpha,
                                                     descrA,
                                                     nnzA,
                                                     dcsrValA,
                                                     dcsrRowPtrA,
                                                     dcsrColIndA,
                                                     &beta,
                                                     descrB,
                                                     nnzB,
                                                     dcsrValB,
                                                     dcsrRowPtrB,
                                                     dcsrColIndB,
                                                     descrC,
                                                     NULL,
                                                     dcsrRowPtrC,
                                                     NULL,
                                                     &bufferSize));

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

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgeam2Nnz(handle,
                                          m,
                                          n,
                                          descrA,
                                          nnzA,
                                          dcsrRowPtrA,
                                          dcsrColIndA,
                                          descrB,
                                          nnzB,
                                          dcsrRowPtrB,
                                          dcsrColIndB,
                                          descrC,
                                          dcsrRowPtrC,
                                          &nnzC,
                                          dbuffer));

    int*   dcsrColIndC = NULL;
    float* dcsrValC    = NULL;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgeam2(handle,
                                       m,
                                       n,
                                       &alpha,
                                       descrA,
                                       nnzA,
                                       dcsrValA,
                                       dcsrRowPtrA,
                                       dcsrColIndA,
                                       &beta,
                                       descrB,
                                       nnzB,
                                       dcsrValB,
                                       dcsrRowPtrB,
                                       dcsrColIndB,
                                       descrC,
                                       dcsrValC,
                                       dcsrRowPtrC,
                                       dcsrColIndC,
                                       dbuffer));

    int*  hcsrRowPtrC = (int*)malloc((m + 1) * sizeof(int));
    int*  hcsrColIndC = (int*)malloc((nnzC) * sizeof(int));
    float hcsrValC[nnzC];

    // Copy back to the host
    HIP_CHECK(hipMemcpy(hcsrRowPtrC, dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrColIndC, dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC, dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    printf("C\n");
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        float* temp = (float*)malloc(n * sizeof(float));
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

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

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIP_CHECK(hipFree(dbuffer));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

program example_fortran_csrgeam2
    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_a(5), h_csr_col_ind_a(9)
    integer, target :: h_csr_row_ptr_b(5), h_csr_col_ind_b(6)
    integer, target :: h_csr_row_ptr_c(5)
    integer, allocatable, target :: h_csr_col_ind_c(:)
    real(4), target :: h_csr_val_a(9), h_csr_val_b(6)
    real(4), allocatable, target :: h_csr_val_c(:)
    real(4) :: temp(4)

    type(c_ptr) :: d_csr_row_ptr_a, d_csr_col_ind_a, d_csr_val_a
    type(c_ptr) :: d_csr_row_ptr_b, d_csr_col_ind_b, d_csr_val_b
    type(c_ptr) :: d_csr_row_ptr_c, d_csr_col_ind_c, d_csr_val_c
    type(c_ptr) :: d_buffer

    integer :: i, j, row_start, row_end
    integer(c_int) :: m, n, nnz_a, nnz_b
    integer(c_int), target :: nnz_c
    integer(c_size_t), target :: buffer_size

    real(c_float), target :: alpha, beta

    type(c_ptr) :: handle
    type(c_ptr) :: descr_a, descr_b, descr_c

!   Input data
    m = 4
    n = 4
    nnz_a = 9
    nnz_b = 6

!   Matrix A (4x4)
    h_csr_row_ptr_a = (/0, 2, 4, 8, 9/)
    h_csr_col_ind_a = (/0, 3, 0, 1, 0, 1, 2, 3, 2/)
    h_csr_val_a     = (/1.0, 2.0, 3.0, 4.0, 5.0, 6.0, 7.0, 8.0, 9.0/)

!   Matrix B (4x4)
    h_csr_row_ptr_b = (/0, 1, 3, 5, 6/)
    h_csr_col_ind_b = (/1, 0, 2, 1, 3, 2/)
    h_csr_val_b     = (/1.0, 1.0, 1.0, 1.0, 1.0, 1.0/)

!   Scalar alpha and beta
    alpha = 1.0
    beta  = 1.0

!   Allocate device memory for A and B
    call HIP_CHECK(hipMalloc(d_csr_row_ptr_a, int(m + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_col_ind_a, int(nnz_a, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_val_a, int(nnz_a, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_row_ptr_b, int(m + 1, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_col_ind_b, int(nnz_b, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_val_b, int(nnz_b, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_row_ptr_c, int(m + 1, c_size_t) * 4))

!   Copy host data to device
    call HIP_CHECK(hipMemcpy(d_csr_row_ptr_a, c_loc(h_csr_row_ptr_a), int(m + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_col_ind_a, c_loc(h_csr_col_ind_a), int(nnz_a, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_val_a, c_loc(h_csr_val_a), int(nnz_a, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_row_ptr_b, c_loc(h_csr_row_ptr_b), int(m + 1, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_col_ind_b, c_loc(h_csr_col_ind_b), int(nnz_b, c_size_t) * 4, 1))
    call HIP_CHECK(hipMemcpy(d_csr_val_b, c_loc(h_csr_val_b), int(nnz_b, c_size_t) * 4, 1))

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

!   Create matrix descriptors
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr_a))
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr_b))
    call HIPSPARSE_CHECK(hipsparseCreateMatDescr(descr_c))

!   Get buffer size
    call HIPSPARSE_CHECK(hipsparseScsrgeam2_bufferSizeExt(handle, &
                                                          m, &
                                                          n, &
                                                          c_loc(alpha), &
                                                          descr_a, &
                                                          nnz_a, &
                                                          d_csr_val_a, &
                                                          d_csr_row_ptr_a, &
                                                          d_csr_col_ind_a, &
                                                          c_loc(beta), &
                                                          descr_b, &
                                                          nnz_b, &
                                                          d_csr_val_b, &
                                                          d_csr_row_ptr_b, &
                                                          d_csr_col_ind_b, &
                                                          descr_c, &
                                                          c_null_ptr, &
                                                          d_csr_row_ptr_c, &
                                                          c_null_ptr, &
                                                          c_loc(buffer_size)))

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

!   Get nnz of C
    call HIPSPARSE_CHECK(hipsparseXcsrgeam2Nnz(handle, &
                                               m, &
                                               n, &
                                               descr_a, &
                                               nnz_a, &
                                               d_csr_row_ptr_a, &
                                               d_csr_col_ind_a, &
                                               descr_b, &
                                               nnz_b, &
                                               d_csr_row_ptr_b, &
                                               d_csr_col_ind_b, &
                                               descr_c, &
                                               d_csr_row_ptr_c, &
                                               c_loc(nnz_c), &
                                               d_buffer))

!   Allocate device memory for C
    call HIP_CHECK(hipMalloc(d_csr_col_ind_c, int(nnz_c, c_size_t) * 4))
    call HIP_CHECK(hipMalloc(d_csr_val_c, int(nnz_c, c_size_t) * 4))

!   Call scsrgeam2 to compute C = alpha * A + beta * B
    call HIPSPARSE_CHECK(hipsparseScsrgeam2(handle, &
                                           m, &
                                           n, &
                                           c_loc(alpha), &
                                           descr_a, &
                                           nnz_a, &
                                           d_csr_val_a, &
                                           d_csr_row_ptr_a, &
                                           d_csr_col_ind_a, &
                                           c_loc(beta), &
                                           descr_b, &
                                           nnz_b, &
                                           d_csr_val_b, &
                                           d_csr_row_ptr_b, &
                                           d_csr_col_ind_b, &
                                           descr_c, &
                                           d_csr_val_c, &
                                           d_csr_row_ptr_c, &
                                           d_csr_col_ind_c, &
                                           d_buffer))

!   Allocate host memory for C
    allocate(h_csr_col_ind_c(nnz_c))
    allocate(h_csr_val_c(nnz_c))

!   Copy result back to host
    call HIP_CHECK(hipMemcpy(c_loc(h_csr_row_ptr_c), d_csr_row_ptr_c, int(m + 1, c_size_t) * 4, 2))
    call HIP_CHECK(hipMemcpy(c_loc(h_csr_col_ind_c), d_csr_col_ind_c, int(nnz_c, c_size_t) * 4, 2))
    call HIP_CHECK(hipMemcpy(c_loc(h_csr_val_c), d_csr_val_c, int(nnz_c, c_size_t) * 4, 2))

!   Print result in dense format
    write(*,'(A)') 'C'
    do i = 1, m
        row_start = h_csr_row_ptr_c(i) + 1
        row_end = h_csr_row_ptr_c(i + 1)
        
        ! Initialize temp array to zeros
        temp = 0.0
        
        ! Fill in non-zero values
        do j = row_start, row_end
            temp(h_csr_col_ind_c(j) + 1) = h_csr_val_c(j)
        end do
        
        ! Print the row
        do j = 1, n
            write(*,'(I0,1X)',advance='no') int(temp(j))
        end do
        write(*,*)
    end do
    write(*,*)

!   Deallocate host arrays
    deallocate(h_csr_col_ind_c)
    deallocate(h_csr_val_c)

!   Clear hipSPARSE
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr_a))
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr_b))
    call HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descr_c))
    call HIPSPARSE_CHECK(hipsparseDestroy(handle))

!   Clear device memory
    call HIP_CHECK(hipFree(d_csr_row_ptr_a))
    call HIP_CHECK(hipFree(d_csr_col_ind_a))
    call HIP_CHECK(hipFree(d_csr_val_a))
    call HIP_CHECK(hipFree(d_csr_row_ptr_b))
    call HIP_CHECK(hipFree(d_csr_col_ind_b))
    call HIP_CHECK(hipFree(d_csr_val_b))
    call HIP_CHECK(hipFree(d_csr_row_ptr_c))
    call HIP_CHECK(hipFree(d_csr_col_ind_c))
    call HIP_CHECK(hipFree(d_csr_val_c))
    call HIP_CHECK(hipFree(d_buffer))

end program example_fortran_csrgeam2

hipsparseXcsrgemmNnz()#

hipsparseStatus_t hipsparseXcsrgemmNnz(hipsparseHandle_t handle, hipsparseOperation_t transA, hipsparseOperation_t transB, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr)#

hipsparseXcsrgemmNnz computes the total CSR non-zero elements and the CSR row offsets, that point to the start of every row of the sparse CSR matrix, of the resulting multiplied matrix \(C\). It is assumed that csrRowPtrC has been allocated with size m+1. The desired index base in the output CSR matrix \(C\) is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

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

As indicated, nnzTotalDevHostPtr can point either to host or device memory. This is controlled by setting the pointer mode. See hipsparseSetPointerMode().

Note

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

Note

Please note, that for matrix products with more than 8192 intermediate products per row, additional temporary storage buffer is allocated by the algorithm.

Note

Currently, only transA == transB == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix \(A\) operation type.
transB – [in] matrix \(B\) operation type.
m – [in] number of rows of the sparse CSR matrix \(op(A)\) and \(C\). Must be non-negative.
n – [in] number of columns of the sparse CSR matrix \(op(B)\) and \(C\). Must be non-negative.
k – [in] number of columns of the sparse CSR matrix \(op(A)\) and number of rows of the sparse CSR matrix \(op(B)\). Must be non-negative.
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\). Must be non-negative.
csrRowPtrA – [in] array of m+1 elements ( \(op(A) == A\), k+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(A)\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\). Must be non-negative.
csrRowPtrB – [in] array of k+1 elements ( \(op(B) == B\), m+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(B)\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
nnzTotalDevHostPtr – [inout] pointer to the number of non-zero entries of the sparse CSR matrix \(C\). nnzTotalDevHostPtr can be a host or device pointer.

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_NOT_INITIALIZED – handle is not initialized.
HIPSPARSE_STATUS_INVALID_VALUE – handle, descrA, descrB or descrC is nullptr, m, n, k, nnzA or nnzB is negative, or csrRowPtrA, csrColIndA, csrRowPtrB, csrColIndB, csrRowPtrC or nnzTotalDevHostPtr is nullptr.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA is not HIPSPARSE_OPERATION_NON_TRANSPOSE, transB is not HIPSPARSE_OPERATION_NON_TRANSPOSE, or hipsparseMatrixType_t is not HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgemm()#

hipsparseStatus_t hipsparseScsrgemm(hipsparseHandle_t handle, hipsparseOperation_t transA, hipsparseOperation_t transB, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const float *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, float *csrValC, const int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseDcsrgemm(hipsparseHandle_t handle, hipsparseOperation_t transA, hipsparseOperation_t transB, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const double *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, double *csrValC, const int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseCcsrgemm(hipsparseHandle_t handle, hipsparseOperation_t transA, hipsparseOperation_t transB, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const hipComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const hipComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, hipComplex *csrValC, const int *csrRowPtrC, int *csrColIndC)#

hipsparseStatus_t hipsparseZcsrgemm(hipsparseHandle_t handle, hipsparseOperation_t transA, hipsparseOperation_t transB, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const hipDoubleComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const hipDoubleComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrValC, const int *csrRowPtrC, int *csrColIndC)#

Sparse matrix sparse matrix multiplication using CSR storage format.

hipsparseXcsrgemm multiplies the sparse \(m \times k\) matrix \(op(A)\), defined in CSR storage format with the sparse \(k \times n\) matrix \(op(B)\), defined in CSR storage format, and stores the result in the sparse \(m \times n\) matrix \(C\), defined in CSR storage format, such that

\[ C := op(A) \cdot op(B), \]

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}\]

and

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

This computation involves a multi step process. First the user must allocate csrRowPtrC to have size m+1. The user then calls hipsparseXcsrgemmNnz which fills in the csrRowPtrC array as well as computes the total number of nonzeros in C, nnzC. The user then allocates both arrays csrColIndC and csrValC to have size nnzC and calls hipsparseXcsrgemm to complete the computation. The desired index base in the output CSR matrix C is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

Currently, only transA == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Note

Currently, only transB == HIPSPARSE_OPERATION_NON_TRANSPOSE is supported.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Note

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

Note

Please note, that for matrix products with more than 4096 non-zero entries per row, additional temporary storage buffer is allocated by the algorithm.

Parameters:

handle – [in] handle to the hipsparse library context queue.
transA – [in] matrix \(A\) operation type.
transB – [in] matrix \(B\) operation type.
m – [in] number of rows of the sparse CSR matrix \(op(A)\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(op(B)\) and \(C\).
k – [in] number of columns of the sparse CSR matrix \(op(A)\) and number of rows of the sparse CSR matrix \(op(B)\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrValA – [in] array of nnzA elements of the sparse CSR matrix \(A\).
csrRowPtrA – [in] array of m+1 elements ( \(op(A) == A\), k+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(A)\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrValB – [in] array of nnzB elements of the sparse CSR matrix \(B\).
csrRowPtrB – [in] array of k+1 elements ( \(op(B) == B\), m+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(B)\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrValC – [out] array of nnzC elements of the sparse CSR matrix \(C\).
csrRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
csrColIndC – [out] array of nnzC elements containing the column indices of the sparse CSR matrix \(C\).

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, k, nnzA, nnzB, descrA, csrValA, csrRowPtrA, csrColIndA, descrB, csrValB, csrRowPtrB, csrColIndB, descrC, csrValC, csrRowPtrC, csrColIndC is invalid.
HIPSPARSE_STATUS_ALLOC_FAILED – additional buffer for long rows could not be allocated.
HIPSPARSE_STATUS_NOT_SUPPORTED – transA != HIPSPARSE_OPERATION_NON_TRANSPOSE, transB != HIPSPARSE_OPERATION_NON_TRANSPOSE, or hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int k    = 3;
    const int n    = 2;
    const int nnzA = 7;
    const int nnzB = 3;

    hipsparseOperation_t transA = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseOperation_t transB = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // A, B, and C are mxk, kxn, and m×n

    // A
    // 1 0 0
    // 3 4 0
    // 5 6 7
    // 0 0 9
    std::vector<int>   hcsrRowPtrA = {0, 1, 3, 6, 7};
    std::vector<int>   hcsrColIndA = {0, 0, 1, 0, 1, 2, 2};
    std::vector<float> hcsrValA    = {1.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 9.0f};

    // B
    // 0 1
    // 1 0
    // 0 1
    std::vector<int>   hcsrRowPtrB = {0, 1, 2, 3};
    std::vector<int>   hcsrColIndB = {1, 0, 1};
    std::vector<float> hcsrValB    = {1.0f, 1.0f, 1.0f};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtrA, hcsrRowPtrA.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndA, hcsrColIndA.data(), nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA.data(), nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrRowPtrB, hcsrRowPtrB.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndB, hcsrColIndB.data(), nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB.data(), nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgemmNnz(handle,
                                         transA,
                                         transB,
                                         m,
                                         n,
                                         k,
                                         descrA,
                                         nnzA,
                                         dcsrRowPtrA,
                                         dcsrColIndA,
                                         descrB,
                                         nnzB,
                                         dcsrRowPtrB,
                                         dcsrColIndB,
                                         descrC,
                                         dcsrRowPtrC,
                                         &nnzC));

    int*   dcsrColIndC = nullptr;
    float* dcsrValC    = nullptr;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgemm(handle,
                                      transA,
                                      transB,
                                      m,
                                      n,
                                      k,
                                      descrA,
                                      nnzA,
                                      dcsrValA,
                                      dcsrRowPtrA,
                                      dcsrColIndA,
                                      descrB,
                                      nnzB,
                                      dcsrValB,
                                      dcsrRowPtrB,
                                      dcsrColIndB,
                                      descrC,
                                      dcsrValC,
                                      dcsrRowPtrC,
                                      dcsrColIndC));

    std::vector<int>   hcsrRowPtrC(m + 1);
    std::vector<int>   hcsrColIndC(nnzC);
    std::vector<float> hcsrValC(nnzC);

    // Copy back to the host
    HIP_CHECK(
        hipMemcpy(hcsrRowPtrC.data(), dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(
        hipMemcpy(hcsrColIndC.data(), dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC.data(), dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    std::cout << "C" << std::endl;
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        std::vector<float> temp(n, 0.0f);
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

        for(int j = 0; j < n; j++)
        {
            std::cout << temp[j] << " ";
        }
        std::cout << std::endl;
    }
    std::cout << std::endl;

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

int main(int argc, char* argv[])
{
    const int m    = 4;
    const int k    = 3;
    const int n    = 2;
    const int nnzA = 7;
    const int nnzB = 3;

    hipsparseOperation_t transA = HIPSPARSE_OPERATION_NON_TRANSPOSE;
    hipsparseOperation_t transB = HIPSPARSE_OPERATION_NON_TRANSPOSE;

    // A, B, and C are mxk, kxn, and m×n

    // A
    // 1 0 0
    // 3 4 0
    // 5 6 7
    // 0 0 9
    int   hcsrRowPtrA[] = {0, 1, 3, 6, 7};
    int   hcsrColIndA[] = {0, 0, 1, 0, 1, 2, 2};
    float hcsrValA[]    = {1.0, 3.0, 4.0, 5.0, 6.0, 7.0, 9.0};

    // B
    // 0 1
    // 1 0
    // 0 1
    int   hcsrRowPtrB[] = {0, 1, 2, 3};
    int   hcsrColIndB[] = {1, 0, 1};
    float hcsrValB[]    = {1.0, 1.0, 1.0};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(hipMemcpy(dcsrRowPtrA, hcsrRowPtrA, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndA, hcsrColIndA, nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA, nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrRowPtrB, hcsrRowPtrB, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndB, hcsrColIndB, nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB, nnzB * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgemmNnz(handle,
                                         transA,
                                         transB,
                                         m,
                                         n,
                                         k,
                                         descrA,
                                         nnzA,
                                         dcsrRowPtrA,
                                         dcsrColIndA,
                                         descrB,
                                         nnzB,
                                         dcsrRowPtrB,
                                         dcsrColIndB,
                                         descrC,
                                         dcsrRowPtrC,
                                         &nnzC));

    int*   dcsrColIndC = NULL;
    float* dcsrValC    = NULL;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgemm(handle,
                                      transA,
                                      transB,
                                      m,
                                      n,
                                      k,
                                      descrA,
                                      nnzA,
                                      dcsrValA,
                                      dcsrRowPtrA,
                                      dcsrColIndA,
                                      descrB,
                                      nnzB,
                                      dcsrValB,
                                      dcsrRowPtrB,
                                      dcsrColIndB,
                                      descrC,
                                      dcsrValC,
                                      dcsrRowPtrC,
                                      dcsrColIndC));

    int*  hcsrRowPtrC = (int*)malloc((m + 1) * sizeof(int));
    int*  hcsrColIndC = (int*)malloc((nnzC) * sizeof(int));
    float hcsrValC[nnzC];

    // Copy back to the host
    HIP_CHECK(hipMemcpy(hcsrRowPtrC, dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrColIndC, dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC, dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    printf("C\n");
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        float* temp = (float*)calloc(n, sizeof(float));
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

        for(int j = 0; j < n; j++)
        {
            printf("%f ", temp[j]);
        }
        printf("\n");
        free(temp);
    }
    printf("\n");

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

program example_hipsparse_csrgemm
    use iso_c_binding
    implicit none

    ! HIP
    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer(c_int) :: 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(c_int) :: 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(c_int) :: 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, parameter :: hipMemcpyHostToDevice = 1
    integer, parameter :: hipMemcpyDeviceToHost = 2

    ! hipSPARSE
    interface
        function hipsparseCreate(handle) &
                bind(c, name = 'hipsparseCreate')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreate
            type(c_ptr) :: handle
        end function hipsparseCreate

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

        function hipsparseCreateMatDescr(descr) &
                bind(c, name = 'hipsparseCreateMatDescr')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreateMatDescr
            type(c_ptr) :: descr
        end function hipsparseCreateMatDescr

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

        function hipsparseXcsrgemmNnz(handle, transA, transB, m, n, k, descrA, nnzA, csrRowPtrA, &
                                      csrColIndA, descrB, nnzB, csrRowPtrB, csrColIndB, descrC, &
                                      csrRowPtrC, nnzTotalDevHostPtr) &
                bind(c, name = 'hipsparseXcsrgemmNnz')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseXcsrgemmNnz
            type(c_ptr), value :: handle
            integer(c_int), value :: transA
            integer(c_int), value :: transB
            integer(c_int), value :: m
            integer(c_int), value :: n
            integer(c_int), value :: k
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrRowPtrA
            type(c_ptr), intent(in), value :: csrColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrRowPtrB
            type(c_ptr), intent(in), value :: csrColIndB
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrRowPtrC
            type(c_ptr), value :: nnzTotalDevHostPtr
        end function hipsparseXcsrgemmNnz

        function hipsparseScsrgemm(handle, transA, transB, m, n, k, descrA, nnzA, csrSortedValA, &
                                   csrSortedRowPtrA, csrSortedColIndA, descrB, nnzB, csrSortedValB, &
                                   csrSortedRowPtrB, csrSortedColIndB, descrC, csrSortedValC, &
                                   csrSortedRowPtrC, csrSortedColIndC) &
                bind(c, name = 'hipsparseScsrgemm')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseScsrgemm
            type(c_ptr), value :: handle
            integer(c_int), value :: transA
            integer(c_int), value :: transB
            integer(c_int), value :: m
            integer(c_int), value :: n
            integer(c_int), value :: k
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrSortedValA
            type(c_ptr), intent(in), value :: csrSortedRowPtrA
            type(c_ptr), intent(in), value :: csrSortedColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrSortedValB
            type(c_ptr), intent(in), value :: csrSortedRowPtrB
            type(c_ptr), intent(in), value :: csrSortedColIndB
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrSortedValC
            type(c_ptr), value :: csrSortedRowPtrC
            type(c_ptr), value :: csrSortedColIndC
        end function hipsparseScsrgemm
    end interface

    integer, parameter :: HIPSPARSE_OPERATION_NON_TRANSPOSE = 0

    ! Variables
    type(c_ptr) :: handle
    type(c_ptr) :: descrA, descrB, descrC
    integer :: i, stat
    
    ! Sparse matrix-matrix multiply: C = op(A) * op(B)
    integer, parameter :: m = 4
    integer, parameter :: k = 3
    integer, parameter :: n = 2
    integer, parameter :: nnzA = 7
    integer, parameter :: nnzB = 3
    integer(c_int), target :: nnzC
    
    ! Matrix A (4x3)
    integer, dimension(m+1), target :: hcsrRowPtrA = (/0, 1, 3, 6, 7/)
    integer, dimension(nnzA), target :: hcsrColIndA = (/0, 0, 1, 0, 1, 2, 2/)
    real(c_float), dimension(nnzA), target :: hcsrValA = (/1.0, 3.0, 4.0, 5.0, 6.0, 7.0, 9.0/)
    
    ! Matrix B (3x2)
    integer, dimension(k+1), target :: hcsrRowPtrB = (/0, 1, 2, 3/)
    integer, dimension(nnzB), target :: hcsrColIndB = (/1, 0, 1/)
    real(c_float), dimension(nnzB), target :: hcsrValB = (/1.0, 1.0, 1.0/)
    
    ! Matrix C (will be allocated after nnzC is determined)
    integer, dimension(:), allocatable, target :: hcsrRowPtrC
    integer, dimension(:), allocatable, target :: hcsrColIndC
    real(c_float), dimension(:), allocatable, target :: hcsrValC
    
    ! Device pointers
    type(c_ptr) :: dcsrRowPtrA, dcsrColIndA, dcsrValA
    type(c_ptr) :: dcsrRowPtrB, dcsrColIndB, dcsrValB
    type(c_ptr) :: dcsrRowPtrC, dcsrColIndC, dcsrValC

    ! Create hipSPARSE handle
    stat = hipsparseCreate(handle)
    if (stat /= 0) stop

    ! Create matrix descriptors
    stat = hipsparseCreateMatDescr(descrA)
    if (stat /= 0) stop
    stat = hipsparseCreateMatDescr(descrB)
    if (stat /= 0) stop
    stat = hipsparseCreateMatDescr(descrC)
    if (stat /= 0) stop

    ! Allocate device memory for A and B
    stat = hipMalloc(dcsrRowPtrA, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrB, int((k + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrC, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop

    ! Copy A and B to device
    stat = hipMemcpy(dcsrRowPtrA, c_loc(hcsrRowPtrA), int((m + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndA, c_loc(hcsrColIndA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValA, c_loc(hcsrValA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    stat = hipMemcpy(dcsrRowPtrB, c_loc(hcsrRowPtrB), int((k + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndB, c_loc(hcsrColIndB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValB, c_loc(hcsrValB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    ! Determine nnzC
    stat = hipsparseXcsrgemmNnz(handle, &
                                HIPSPARSE_OPERATION_NON_TRANSPOSE, &
                                HIPSPARSE_OPERATION_NON_TRANSPOSE, &
                                m, &
                                n, &
                                k, &
                                descrA, &
                                nnzA, &
                                dcsrRowPtrA, &
                                dcsrColIndA, &
                                descrB, &
                                nnzB, &
                                dcsrRowPtrB, &
                                dcsrColIndB, &
                                descrC, &
                                dcsrRowPtrC, &
                                c_loc(nnzC))
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseXcsrgemmNnz failed'
        stop
    end if

    ! Allocate device memory for C
    stat = hipMalloc(dcsrColIndC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop

    ! Perform matrix multiplication: C = A * B
    stat = hipsparseScsrgemm(handle, &
                             HIPSPARSE_OPERATION_NON_TRANSPOSE, &
                             HIPSPARSE_OPERATION_NON_TRANSPOSE, &
                             m, &
                             n, &
                             k, &
                             descrA, &
                             nnzA, &
                             dcsrValA, &
                             dcsrRowPtrA, &
                             dcsrColIndA, &
                             descrB, &
                             nnzB, &
                             dcsrValB, &
                             dcsrRowPtrB, &
                             dcsrColIndB, &
                             descrC, &
                             dcsrValC, &
                             dcsrRowPtrC, &
                             dcsrColIndC)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseScsrgemm failed'
        stop
    end if

    ! Allocate host memory for C
    allocate(hcsrRowPtrC(m+1))
    allocate(hcsrColIndC(nnzC))
    allocate(hcsrValC(nnzC))

    ! Copy result back to host
    stat = hipMemcpy(c_loc(hcsrRowPtrC), dcsrRowPtrC, int((m + 1) * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrColIndC), dcsrColIndC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrValC), dcsrValC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop

    ! Print result
    write(*,*) 'Matrix C (result of A * B):'
    write(*,*) 'nnzC =', nnzC
    write(*,*) 'csrRowPtrC:'
    do i = 1, m + 1
        write(*,*) hcsrRowPtrC(i)
    end do
    write(*,*) 'csrColIndC:'
    do i = 1, nnzC
        write(*,*) hcsrColIndC(i)
    end do
    write(*,*) 'csrValC:'
    do i = 1, nnzC
        write(*,*) hcsrValC(i)
    end do

    ! Clean up
    deallocate(hcsrRowPtrC)
    deallocate(hcsrColIndC)
    deallocate(hcsrValC)

    stat = hipFree(dcsrRowPtrA)
    stat = hipFree(dcsrColIndA)
    stat = hipFree(dcsrValA)
    stat = hipFree(dcsrRowPtrB)
    stat = hipFree(dcsrColIndB)
    stat = hipFree(dcsrValB)
    stat = hipFree(dcsrRowPtrC)
    stat = hipFree(dcsrColIndC)
    stat = hipFree(dcsrValC)

    stat = hipsparseDestroyMatDescr(descrA)
    stat = hipsparseDestroyMatDescr(descrB)
    stat = hipsparseDestroyMatDescr(descrC)
    stat = hipsparseDestroy(handle)

end program example_hipsparse_csrgemm

hipsparseXcsrgemm2_bufferSizeExt()#

hipsparseStatus_t hipsparseScsrgemm2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int k, const float *alpha, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const float *beta, const hipsparseMatDescr_t descrD, int nnzD, const int *csrRowPtrD, const int *csrColIndD, csrgemm2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseDcsrgemm2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int k, const double *alpha, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const double *beta, const hipsparseMatDescr_t descrD, int nnzD, const int *csrRowPtrD, const int *csrColIndD, csrgemm2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseCcsrgemm2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int k, const hipComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const hipComplex *beta, const hipsparseMatDescr_t descrD, int nnzD, const int *csrRowPtrD, const int *csrColIndD, csrgemm2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseStatus_t hipsparseZcsrgemm2_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int k, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const hipDoubleComplex *beta, const hipsparseMatDescr_t descrD, int nnzD, const int *csrRowPtrD, const int *csrColIndD, csrgemm2Info_t info, size_t *pBufferSizeInBytes)#

hipsparseXcsrgemm2_bufferSizeExt returns the size of the temporary storage buffer in bytes that is required by hipsparseXcsrgemm2Nnz() and hipsparseXcsrgemm2(). The temporary storage buffer must be allocated by the user.

Note

Please note, that for matrix products with more than 4096 non-zero entries per row, additional temporary storage buffer is allocated by the algorithm.

Note

Please note, that for matrix products with more than 8192 intermediate products per row, additional temporary storage buffer is allocated by the algorithm.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(op(A)\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(op(B)\) and \(C\).
k – [in] number of columns of the sparse CSR matrix \(op(A)\) and number of rows of the sparse CSR matrix \(op(B)\).
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrRowPtrA – [in] array of m+1 elements ( \(op(A) == A\), k+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(A)\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrRowPtrB – [in] array of k+1 elements ( \(op(B) == B\), m+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(B)\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
beta – [in] scalar \(\beta\).
descrD – [in] descriptor of the sparse CSR matrix \(D\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzD – [in] number of non-zero entries of the sparse CSR matrix \(D\).
csrRowPtrD – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(D\).
csrColIndD – [in] array of nnzD elements containing the column indices of the sparse CSR matrix \(D\).
info – [inout] structure that holds meta data for the sparse CSR matrix \(C\).
pBufferSizeInBytes – [out] number of bytes of the temporary storage buffer required by hipsparseXcsrgemm2Nnz(), hipsparseScsrgemm2(), hipsparseDcsrgemm2(), hipsparseCcsrgemm2() and hipsparseZcsrgemm2().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, k, nnzA, nnzB, nnz_D, alpha, beta, descrA, csrRowPtrA, csrColIndA, descrB, csrRowPtrB, csrColIndB, descrD, csrRowPtrD, csrColIndD, info or pBufferSizeInBytes is invalid.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgemm2Nnz()#

hipsparseStatus_t hipsparseXcsrgemm2Nnz(hipsparseHandle_t handle, int m, int n, int k, const hipsparseMatDescr_t descrA, int nnzA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const int *csrRowPtrB, const int *csrColIndB, const hipsparseMatDescr_t descrD, int nnzD, const int *csrRowPtrD, const int *csrColIndD, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr, const csrgemm2Info_t info, void *pBuffer)#

hipsparseXcsrgemm2Nnz computes the total CSR non-zero elements and the CSR row offsets, that point to the start of every row of the sparse CSR matrix, of the resulting multiplied matrix \(C\). It is assumed that csrRowPtrC has been allocated with size m+1. The required buffer size can be obtained by hipsparseXcsrgemm2_bufferSizeExt(). The desired index base in the output CSR matrix \(C\) is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

As indicated, nnzTotalDevHostPtr can point either to host or device memory. This is controlled by setting the pointer mode. See hipsparseSetPointerMode().

Note

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

Note

Please note, that for matrix products with more than 8192 intermediate products per row, additional temporary storage buffer is allocated by the algorithm.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(op(A)\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(op(B)\) and \(C\).
k – [in] number of columns of the sparse CSR matrix \(op(A)\) and number of rows of the sparse CSR matrix \(op(B)\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrRowPtrA – [in] array of m+1 elements ( \(op(A) == A\), k+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(A)\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrRowPtrB – [in] array of k+1 elements ( \(op(B) == B\), m+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(B)\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
descrD – [in] descriptor of the sparse CSR matrix \(D\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzD – [in] number of non-zero entries of the sparse CSR matrix \(D\).
csrRowPtrD – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(D\).
csrColIndD – [in] array of nnzD elements containing the column indices of the sparse CSR matrix \(D\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrRowPtrC – [out] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
nnzTotalDevHostPtr – [out] pointer to the number of non-zero entries of the sparse CSR matrix \(C\).
info – [in] structure that holds meta data for the sparse CSR matrix \(C\).
pBuffer – [in] temporary storage buffer allocated by the user, size is returned by hipsparseScsrgemm2_bufferSizeExt(), hipsparseDcsrgemm2_bufferSizeExt(), hipsparseZcsrgemm2_bufferSizeExt() or hipsparseZcsrgemm2_bufferSizeExt().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, k, nnzA, nnzB, nnzD, descrA, csrRowPtrA, csrColIndA, descrB, csrRowPtrB, csrColIndB, descrD, csrRowPtrD, csrColIndD, descrC, csrRowPtrC, nnzTotalDevHostPtr, info or pBuffer is invalid.
HIPSPARSE_STATUS_ALLOC_FAILED – additional buffer for long rows could not be allocated.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

hipsparseXcsrgemm2()#

hipsparseStatus_t hipsparseScsrgemm2(hipsparseHandle_t handle, int m, int n, int k, const float *alpha, const hipsparseMatDescr_t descrA, int nnzA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const float *csrValB, const int *csrRowPtrB, const int *csrColIndB, const float *beta, const hipsparseMatDescr_t descrD, int nnzD, const float *csrValD, const int *csrRowPtrD, const int *csrColIndD, const hipsparseMatDescr_t descrC, float *csrValC, const int *csrRowPtrC, int *csrColIndC, const csrgemm2Info_t info, void *pBuffer)#

hipsparseStatus_t hipsparseDcsrgemm2(hipsparseHandle_t handle, int m, int n, int k, const double *alpha, const hipsparseMatDescr_t descrA, int nnzA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const double *csrValB, const int *csrRowPtrB, const int *csrColIndB, const double *beta, const hipsparseMatDescr_t descrD, int nnzD, const double *csrValD, const int *csrRowPtrD, const int *csrColIndD, const hipsparseMatDescr_t descrC, double *csrValC, const int *csrRowPtrC, int *csrColIndC, const csrgemm2Info_t info, void *pBuffer)#

hipsparseStatus_t hipsparseCcsrgemm2(hipsparseHandle_t handle, int m, int n, int k, const hipComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const hipComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipComplex *beta, const hipsparseMatDescr_t descrD, int nnzD, const hipComplex *csrValD, const int *csrRowPtrD, const int *csrColIndD, const hipsparseMatDescr_t descrC, hipComplex *csrValC, const int *csrRowPtrC, int *csrColIndC, const csrgemm2Info_t info, void *pBuffer)#

hipsparseStatus_t hipsparseZcsrgemm2(hipsparseHandle_t handle, int m, int n, int k, const hipDoubleComplex *alpha, const hipsparseMatDescr_t descrA, int nnzA, const hipDoubleComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, const hipsparseMatDescr_t descrB, int nnzB, const hipDoubleComplex *csrValB, const int *csrRowPtrB, const int *csrColIndB, const hipDoubleComplex *beta, const hipsparseMatDescr_t descrD, int nnzD, const hipDoubleComplex *csrValD, const int *csrRowPtrD, const int *csrColIndD, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrValC, const int *csrRowPtrC, int *csrColIndC, const csrgemm2Info_t info, void *pBuffer)#

Sparse matrix sparse matrix multiplication using CSR storage format.

hipsparseXcsrgemm2 multiplies the scalar \(\alpha\) with the sparse \(m \times k\) matrix \(A\), defined in CSR storage format, and the sparse \(k \times n\) matrix \(B\), defined in CSR storage format, and adds the result to the sparse \(m \times n\) matrix \(D\) that is multiplied by \(\beta\). The final result is stored in the sparse \(m \times n\) matrix \(C\), defined in CSR storage format, such that

\[ C := \alpha \cdot A \cdot B + \beta \cdot D \]

This computation involves a multi step process. First the user must call hipsparseXcsrgemm2_bufferSizeExt() in order to determine the required user allocated temporary buffer size. The user then allocates this buffer and also allocates csrRowPtrC to have size m+1. Both the temporary storage buffer and csrRowPtrC array are then passed to hipsparseXcsrgemm2Nnz which fills in the csrRowPtrC array as well as computes the total number of nonzeros in C, nnzC. The user then allocates both arrays csrColIndC and csrValC to have size nnzC and calls hipsparseXcsrgemm2 to complete the computation. The desired index base in the output CSR matrix C is set in the hipsparseMatDescr_t descrC. See hipsparseSetMatIndexBase().

Note

If \(\alpha == 0\), then \(C = \beta \cdot D\) will be computed.

Note

If \(\beta == 0\), then \(C = \alpha \cdot A \cdot B\) will be computed.

Note

\(\alpha == beta == 0\) is invalid.

Note

Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.

Note

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

Note

Please note, that for matrix products with more than 4096 non-zero entries per row, additional temporary storage buffer is allocated by the algorithm.

Parameters:

handle – [in] handle to the hipsparse library context queue.
m – [in] number of rows of the sparse CSR matrix \(op(A)\) and \(C\).
n – [in] number of columns of the sparse CSR matrix \(op(B)\) and \(C\).
k – [in] number of columns of the sparse CSR matrix \(op(A)\) and number of rows of the sparse CSR matrix \(op(B)\).
alpha – [in] scalar \(\alpha\).
descrA – [in] descriptor of the sparse CSR matrix \(A\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzA – [in] number of non-zero entries of the sparse CSR matrix \(A\).
csrValA – [in] array of nnzA elements of the sparse CSR matrix \(A\).
csrRowPtrA – [in] array of m+1 elements ( \(op(A) == A\), k+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(A)\).
csrColIndA – [in] array of nnzA elements containing the column indices of the sparse CSR matrix \(A\).
descrB – [in] descriptor of the sparse CSR matrix \(B\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzB – [in] number of non-zero entries of the sparse CSR matrix \(B\).
csrValB – [in] array of nnzB elements of the sparse CSR matrix \(B\).
csrRowPtrB – [in] array of k+1 elements ( \(op(B) == B\), m+1 otherwise) that point to the start of every row of the sparse CSR matrix \(op(B)\).
csrColIndB – [in] array of nnzB elements containing the column indices of the sparse CSR matrix \(B\).
beta – [in] scalar \(\beta\).
descrD – [in] descriptor of the sparse CSR matrix \(D\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
nnzD – [in] number of non-zero entries of the sparse CSR matrix \(D\).
csrValD – [in] array of nnzD elements of the sparse CSR matrix \(D\).
csrRowPtrD – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(D\).
csrColIndD – [in] array of nnzD elements containing the column indices of the sparse CSR matrix \(D\).
descrC – [in] descriptor of the sparse CSR matrix \(C\). Currently, only HIPSPARSE_MATRIX_TYPE_GENERAL is supported.
csrValC – [out] array of nnzC elements of the sparse CSR matrix \(C\).
csrRowPtrC – [in] array of m+1 elements that point to the start of every row of the sparse CSR matrix \(C\).
csrColIndC – [out] array of nnzC elements containing the column indices of the sparse CSR matrix \(C\).
info – [in] structure that holds meta data for the sparse CSR matrix \(C\).
pBuffer – [in] temporary storage buffer allocated by the user, size is returned by hipsparseScsrgemm2_bufferSizeExt(), hipsparseDcsrgemm2_bufferSizeExt(), hipsparseCcsrgemm2_bufferSizeExt() or hipsparseZcsrgemm2_bufferSizeExt().

Return values:

HIPSPARSE_STATUS_SUCCESS – the operation completed successfully.
HIPSPARSE_STATUS_INVALID_VALUE – handle, m, n, k, nnzA, nnzB, nnzD, alpha, beta, descrA, csrValA, csrRowPtrA, csrColIndA, descrB, csrValB, csrRowPtrB, csrColIndB, descrD, csrValD, csrRowPtrD, csrColIndD, csrValC, csrRowPtrC, csrColIndC, info or pBuffer is invalid.
HIPSPARSE_STATUS_ALLOC_FAILED – additional buffer for long rows could not be allocated.
HIPSPARSE_STATUS_NOT_SUPPORTED – hipsparseMatrixType_t != HIPSPARSE_MATRIX_TYPE_GENERAL.

int main(int argc, char* argv[])
{
    int m    = 4;
    int k    = 3;
    int n    = 2;
    int nnzA = 7;
    int nnzB = 3;
    int nnzD = 6;

    float alpha{1.0f};
    float beta{1.0f};

    // A, B, and C are mxk, kxn, and m×n

    // A
    // 1 0 0
    // 3 4 0
    // 5 6 7
    // 0 0 9
    std::vector<int>   hcsrRowPtrA = {0, 1, 3, 6, 7};
    std::vector<int>   hcsrColIndA = {0, 0, 1, 0, 1, 2, 2};
    std::vector<float> hcsrValA    = {1.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 9.0f};

    // B
    // 0 1
    // 1 0
    // 0 1
    std::vector<int>   hcsrRowPtrB = {0, 1, 2, 3};
    std::vector<int>   hcsrColIndB = {1, 0, 1};
    std::vector<float> hcsrValB    = {1.0f, 1.0f, 1.0f};

    // D
    // 0 1
    // 2 3
    // 4 5
    // 0 6
    std::vector<int>   hcsrRowPtrD = {0, 1, 3, 5, 6};
    std::vector<int>   hcsrColIndD = {1, 0, 1, 0, 1, 1};
    std::vector<float> hcsrValD    = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrD;
    int*   dcsrColIndD;
    float* dcsrValD;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (k + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrD, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndD, nnzD * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValD, nnzD * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(
        hipMemcpy(dcsrRowPtrA, hcsrRowPtrA.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndA, hcsrColIndA.data(), nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA.data(), nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrRowPtrB, hcsrRowPtrB.data(), (k + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndB, hcsrColIndB.data(), nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB.data(), nnzB * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrRowPtrD, hcsrRowPtrD.data(), (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(
        hipMemcpy(dcsrColIndD, hcsrColIndD.data(), nnzD * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValD, hcsrValD.data(), nnzD * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    hipsparseMatDescr_t descrD;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrD));

    csrgemm2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateCsrgemm2Info(&info));

    size_t bufferSize;
    HIPSPARSE_CHECK(hipsparseScsrgemm2_bufferSizeExt(handle,
                                                     m,
                                                     n,
                                                     k,
                                                     &alpha,
                                                     descrA,
                                                     nnzA,
                                                     dcsrRowPtrA,
                                                     dcsrColIndA,
                                                     descrB,
                                                     nnzB,
                                                     dcsrRowPtrB,
                                                     dcsrColIndB,
                                                     &beta,
                                                     descrD,
                                                     nnzD,
                                                     dcsrRowPtrD,
                                                     dcsrColIndD,
                                                     info,
                                                     &bufferSize));

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

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgemm2Nnz(handle,
                                          m,
                                          n,
                                          k,
                                          descrA,
                                          nnzA,
                                          dcsrRowPtrA,
                                          dcsrColIndA,
                                          descrB,
                                          nnzB,
                                          dcsrRowPtrB,
                                          dcsrColIndB,
                                          descrD,
                                          nnzD,
                                          dcsrRowPtrD,
                                          dcsrColIndD,
                                          descrC,
                                          dcsrRowPtrC,
                                          &nnzC,
                                          info,
                                          dbuffer));

    int*   dcsrColIndC = nullptr;
    float* dcsrValC    = nullptr;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgemm2(handle,
                                       m,
                                       n,
                                       k,
                                       &alpha,
                                       descrA,
                                       nnzA,
                                       dcsrValA,
                                       dcsrRowPtrA,
                                       dcsrColIndA,
                                       descrB,
                                       nnzB,
                                       dcsrValB,
                                       dcsrRowPtrB,
                                       dcsrColIndB,
                                       &beta,
                                       descrD,
                                       nnzD,
                                       dcsrValD,
                                       dcsrRowPtrD,
                                       dcsrColIndD,
                                       descrC,
                                       dcsrValC,
                                       dcsrRowPtrC,
                                       dcsrColIndC,
                                       info,
                                       dbuffer));

    std::vector<int>   hcsrRowPtrC(m + 1);
    std::vector<int>   hcsrColIndC(nnzC);
    std::vector<float> hcsrValC(nnzC);

    // Copy back to the host
    HIP_CHECK(
        hipMemcpy(hcsrRowPtrC.data(), dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(
        hipMemcpy(hcsrColIndC.data(), dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC.data(), dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    std::cout << "C" << std::endl;
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        std::vector<float> temp(n, 0.0f);
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

        for(int j = 0; j < n; j++)
        {
            std::cout << temp[j] << " ";
        }
        std::cout << std::endl;
    }
    std::cout << std::endl;

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));
    HIP_CHECK(hipFree(dcsrRowPtrD));
    HIP_CHECK(hipFree(dcsrColIndD));
    HIP_CHECK(hipFree(dcsrValD));

    HIP_CHECK(hipFree(dbuffer));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrD));
    HIPSPARSE_CHECK(hipsparseDestroyCsrgemm2Info(info));

    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

int main(int argc, char* argv[])
{
    int m    = 4;
    int k    = 3;
    int n    = 2;
    int nnzA = 7;
    int nnzB = 3;
    int nnzD = 6;

    float alpha = 1.0;
    float beta  = 1.0;

    // A, B, and C are mxk, kxn, and m×n

    // A
    // 1 0 0
    // 3 4 0
    // 5 6 7
    // 0 0 9
    int   hcsrRowPtrA[] = {0, 1, 3, 6, 7};
    int   hcsrColIndA[] = {0, 0, 1, 0, 1, 2, 2};
    float hcsrValA[]    = {1.0, 3.0, 4.0, 5.0, 6.0, 7.0, 9.0};

    // B
    // 0 1
    // 1 0
    // 0 1
    int   hcsrRowPtrB[] = {0, 1, 2, 3};
    int   hcsrColIndB[] = {1, 0, 1};
    float hcsrValB[]    = {1.0, 1.0, 1.0};

    // D
    // 0 1
    // 2 3
    // 4 5
    // 0 6
    int   hcsrRowPtrD[] = {0, 1, 3, 5, 6};
    int   hcsrColIndD[] = {1, 0, 1, 0, 1, 1};
    float hcsrValD[]    = {1.0, 2.0, 3.0, 4.0, 5.0, 6.0};

    // Device memory management: Allocate and copy A, B
    int*   dcsrRowPtrA;
    int*   dcsrColIndA;
    float* dcsrValA;
    int*   dcsrRowPtrB;
    int*   dcsrColIndB;
    float* dcsrValB;
    int*   dcsrRowPtrD;
    int*   dcsrColIndD;
    float* dcsrValD;
    int*   dcsrRowPtrC;
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrA, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndA, nnzA * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValA, nnzA * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrB, (k + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndB, nnzB * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValB, nnzB * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrD, (m + 1) * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrColIndD, nnzD * sizeof(int)));
    HIP_CHECK(hipMalloc((void**)&dcsrValD, nnzD * sizeof(float)));
    HIP_CHECK(hipMalloc((void**)&dcsrRowPtrC, (m + 1) * sizeof(int)));

    HIP_CHECK(hipMemcpy(dcsrRowPtrA, hcsrRowPtrA, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndA, hcsrColIndA, nnzA * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValA, hcsrValA, nnzA * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrRowPtrB, hcsrRowPtrB, (k + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndB, hcsrColIndB, nnzB * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValB, hcsrValB, nnzB * sizeof(float), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrRowPtrD, hcsrRowPtrD, (m + 1) * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrColIndD, hcsrColIndD, nnzD * sizeof(int), hipMemcpyHostToDevice));
    HIP_CHECK(hipMemcpy(dcsrValD, hcsrValD, nnzD * sizeof(float), hipMemcpyHostToDevice));

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

    hipsparseMatDescr_t descrA;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrA));

    hipsparseMatDescr_t descrB;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrB));

    hipsparseMatDescr_t descrC;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrC));

    hipsparseMatDescr_t descrD;
    HIPSPARSE_CHECK(hipsparseCreateMatDescr(&descrD));

    csrgemm2Info_t info;
    HIPSPARSE_CHECK(hipsparseCreateCsrgemm2Info(&info));

    size_t bufferSize;
    HIPSPARSE_CHECK(hipsparseScsrgemm2_bufferSizeExt(handle,
                                                     m,
                                                     n,
                                                     k,
                                                     &alpha,
                                                     descrA,
                                                     nnzA,
                                                     dcsrRowPtrA,
                                                     dcsrColIndA,
                                                     descrB,
                                                     nnzB,
                                                     dcsrRowPtrB,
                                                     dcsrColIndB,
                                                     &beta,
                                                     descrD,
                                                     nnzD,
                                                     dcsrRowPtrD,
                                                     dcsrColIndD,
                                                     info,
                                                     &bufferSize));

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

    int nnzC;
    HIPSPARSE_CHECK(hipsparseXcsrgemm2Nnz(handle,
                                          m,
                                          n,
                                          k,
                                          descrA,
                                          nnzA,
                                          dcsrRowPtrA,
                                          dcsrColIndA,
                                          descrB,
                                          nnzB,
                                          dcsrRowPtrB,
                                          dcsrColIndB,
                                          descrD,
                                          nnzD,
                                          dcsrRowPtrD,
                                          dcsrColIndD,
                                          descrC,
                                          dcsrRowPtrC,
                                          &nnzC,
                                          info,
                                          dbuffer));

    int*   dcsrColIndC = NULL;
    float* dcsrValC    = NULL;
    HIP_CHECK(hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC));
    HIP_CHECK(hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC));

    HIPSPARSE_CHECK(hipsparseScsrgemm2(handle,
                                       m,
                                       n,
                                       k,
                                       &alpha,
                                       descrA,
                                       nnzA,
                                       dcsrValA,
                                       dcsrRowPtrA,
                                       dcsrColIndA,
                                       descrB,
                                       nnzB,
                                       dcsrValB,
                                       dcsrRowPtrB,
                                       dcsrColIndB,
                                       &beta,
                                       descrD,
                                       nnzD,
                                       dcsrValD,
                                       dcsrRowPtrD,
                                       dcsrColIndD,
                                       descrC,
                                       dcsrValC,
                                       dcsrRowPtrC,
                                       dcsrColIndC,
                                       info,
                                       dbuffer));

    int*  hcsrRowPtrC = (int*)malloc((m + 1) * sizeof(int));
    int*  hcsrColIndC = (int*)malloc((nnzC) * sizeof(int));
    float hcsrValC[nnzC];

    // Copy back to the host
    HIP_CHECK(hipMemcpy(hcsrRowPtrC, dcsrRowPtrC, sizeof(int) * (m + 1), hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrColIndC, dcsrColIndC, sizeof(int) * nnzC, hipMemcpyDeviceToHost));
    HIP_CHECK(hipMemcpy(hcsrValC, dcsrValC, sizeof(float) * nnzC, hipMemcpyDeviceToHost));

    printf("C\n");
    for(int i = 0; i < m; i++)
    {
        int start = hcsrRowPtrC[i];
        int end   = hcsrRowPtrC[i + 1];

        float* temp = (float*)malloc(n * sizeof(float));
        for(int j = start; j < end; j++)
        {
            temp[hcsrColIndC[j]] = hcsrValC[j];
        }

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

    HIP_CHECK(hipFree(dcsrRowPtrA));
    HIP_CHECK(hipFree(dcsrColIndA));
    HIP_CHECK(hipFree(dcsrValA));
    HIP_CHECK(hipFree(dcsrRowPtrB));
    HIP_CHECK(hipFree(dcsrColIndB));
    HIP_CHECK(hipFree(dcsrValB));
    HIP_CHECK(hipFree(dcsrRowPtrC));
    HIP_CHECK(hipFree(dcsrColIndC));
    HIP_CHECK(hipFree(dcsrValC));
    HIP_CHECK(hipFree(dcsrRowPtrD));
    HIP_CHECK(hipFree(dcsrColIndD));
    HIP_CHECK(hipFree(dcsrValD));

    HIP_CHECK(hipFree(dbuffer));

    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrA));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrB));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrC));
    HIPSPARSE_CHECK(hipsparseDestroyMatDescr(descrD));
    HIPSPARSE_CHECK(hipsparseDestroyCsrgemm2Info(info));

    HIPSPARSE_CHECK(hipsparseDestroy(handle));

    return 0;
}

program example_hipsparse_csrgemm2
    use iso_c_binding
    implicit none

    ! HIP
    interface
        function hipMalloc(ptr, size) &
                bind(c, name = 'hipMalloc')
            use iso_c_binding
            implicit none
            integer(c_int) :: 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(c_int) :: 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(c_int) :: 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, parameter :: hipMemcpyHostToDevice = 1
    integer, parameter :: hipMemcpyDeviceToHost = 2

    ! hipSPARSE
    interface
        function hipsparseCreate(handle) &
                bind(c, name = 'hipsparseCreate')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreate
            type(c_ptr) :: handle
        end function hipsparseCreate

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

        function hipsparseCreateMatDescr(descr) &
                bind(c, name = 'hipsparseCreateMatDescr')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreateMatDescr
            type(c_ptr) :: descr
        end function hipsparseCreateMatDescr

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

        function hipsparseCreateCsrgemm2Info(info) &
                bind(c, name = 'hipsparseCreateCsrgemm2Info')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseCreateCsrgemm2Info
            type(c_ptr) :: info
        end function hipsparseCreateCsrgemm2Info

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

        function hipsparseScsrgemm2_bufferSizeExt(handle, m, n, k, alpha, descrA, nnzA, csrRowPtrA, &
                                                   csrColIndA, descrB, nnzB, csrRowPtrB, csrColIndB, &
                                                   beta, descrD, nnzD, csrRowPtrD, csrColIndD, &
                                                   info, pBufferSizeInBytes) &
                bind(c, name = 'hipsparseScsrgemm2_bufferSizeExt')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseScsrgemm2_bufferSizeExt
            type(c_ptr), value :: handle
            integer(c_int), value :: m
            integer(c_int), value :: n
            integer(c_int), value :: k
            type(c_ptr), intent(in), value :: alpha
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrRowPtrA
            type(c_ptr), intent(in), value :: csrColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrRowPtrB
            type(c_ptr), intent(in), value :: csrColIndB
            type(c_ptr), intent(in), value :: beta
            type(c_ptr), value :: descrD
            integer(c_int), value :: nnzD
            type(c_ptr), intent(in), value :: csrRowPtrD
            type(c_ptr), intent(in), value :: csrColIndD
            type(c_ptr), value :: info
            type(c_ptr), value :: pBufferSizeInBytes
        end function hipsparseScsrgemm2_bufferSizeExt

        function hipsparseXcsrgemm2Nnz(handle, m, n, k, descrA, nnzA, csrRowPtrA, csrColIndA, &
                                       descrB, nnzB, csrRowPtrB, csrColIndB, descrD, nnzD, &
                                       csrRowPtrD, csrColIndD, descrC, csrRowPtrC, nnzTotalDevHostPtr, &
                                       info, pBuffer) &
                bind(c, name = 'hipsparseXcsrgemm2Nnz')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseXcsrgemm2Nnz
            type(c_ptr), value :: handle
            integer(c_int), value :: m
            integer(c_int), value :: n
            integer(c_int), value :: k
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrRowPtrA
            type(c_ptr), intent(in), value :: csrColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrRowPtrB
            type(c_ptr), intent(in), value :: csrColIndB
            type(c_ptr), value :: descrD
            integer(c_int), value :: nnzD
            type(c_ptr), intent(in), value :: csrRowPtrD
            type(c_ptr), intent(in), value :: csrColIndD
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrRowPtrC
            type(c_ptr), value :: nnzTotalDevHostPtr
            type(c_ptr), value :: info
            type(c_ptr), value :: pBuffer
        end function hipsparseXcsrgemm2Nnz

        function hipsparseScsrgemm2(handle, m, n, k, alpha, descrA, nnzA, csrValA, csrRowPtrA, &
                                    csrColIndA, descrB, nnzB, csrValB, csrRowPtrB, csrColIndB, &
                                    beta, descrD, nnzD, csrValD, csrRowPtrD, csrColIndD, &
                                    descrC, csrValC, csrRowPtrC, csrColIndC, info, pBuffer) &
                bind(c, name = 'hipsparseScsrgemm2')
            use iso_c_binding
            implicit none
            integer(c_int) :: hipsparseScsrgemm2
            type(c_ptr), value :: handle
            integer(c_int), value :: m
            integer(c_int), value :: n
            integer(c_int), value :: k
            type(c_ptr), intent(in), value :: alpha
            type(c_ptr), value :: descrA
            integer(c_int), value :: nnzA
            type(c_ptr), intent(in), value :: csrValA
            type(c_ptr), intent(in), value :: csrRowPtrA
            type(c_ptr), intent(in), value :: csrColIndA
            type(c_ptr), value :: descrB
            integer(c_int), value :: nnzB
            type(c_ptr), intent(in), value :: csrValB
            type(c_ptr), intent(in), value :: csrRowPtrB
            type(c_ptr), intent(in), value :: csrColIndB
            type(c_ptr), intent(in), value :: beta
            type(c_ptr), value :: descrD
            integer(c_int), value :: nnzD
            type(c_ptr), intent(in), value :: csrValD
            type(c_ptr), intent(in), value :: csrRowPtrD
            type(c_ptr), intent(in), value :: csrColIndD
            type(c_ptr), value :: descrC
            type(c_ptr), value :: csrValC
            type(c_ptr), value :: csrRowPtrC
            type(c_ptr), value :: csrColIndC
            type(c_ptr), value :: info
            type(c_ptr), value :: pBuffer
        end function hipsparseScsrgemm2
    end interface

    ! Variables
    type(c_ptr) :: handle
    type(c_ptr) :: descrA, descrB, descrC, descrD
    type(c_ptr) :: info
    integer :: i, stat, start_idx, end_idx
    integer(c_size_t), target :: bufferSize
    
    ! C = alpha * A * B + beta * D
    ! A is m x k, B is k x n, D is m x n, C is m x n
    integer, parameter :: m = 4
    integer, parameter :: k = 3
    integer, parameter :: n = 2
    integer, parameter :: nnzA = 7
    integer, parameter :: nnzB = 3
    integer, parameter :: nnzD = 6
    integer(c_int), target :: nnzC
    
    ! Matrix A (4x3)
    integer, dimension(m+1), target :: hcsrRowPtrA = (/0, 1, 3, 6, 7/)
    integer, dimension(nnzA), target :: hcsrColIndA = (/0, 0, 1, 0, 1, 2, 2/)
    real(c_float), dimension(nnzA), target :: hcsrValA = (/1.0, 3.0, 4.0, 5.0, 6.0, 7.0, 9.0/)
    
    ! Matrix B (3x2)
    integer, dimension(k+1), target :: hcsrRowPtrB = (/0, 1, 2, 3/)
    integer, dimension(nnzB), target :: hcsrColIndB = (/1, 0, 1/)
    real(c_float), dimension(nnzB), target :: hcsrValB = (/1.0, 1.0, 1.0/)
    
    ! Matrix D (4x2)
    integer, dimension(m+1), target :: hcsrRowPtrD = (/0, 1, 3, 5, 6/)
    integer, dimension(nnzD), target :: hcsrColIndD = (/1, 0, 1, 0, 1, 1/)
    real(c_float), dimension(nnzD), target :: hcsrValD = (/1.0, 2.0, 3.0, 4.0, 5.0, 6.0/)
    
    ! Matrix C (will be allocated after nnzC is determined)
    integer, dimension(:), allocatable, target :: hcsrRowPtrC
    integer, dimension(:), allocatable, target :: hcsrColIndC
    real(c_float), dimension(:), allocatable, target :: hcsrValC
    
    ! Scalar values
    real(c_float), target :: alpha = 1.0
    real(c_float), target :: beta = 1.0
    
    ! Device pointers
    type(c_ptr) :: dcsrRowPtrA, dcsrColIndA, dcsrValA
    type(c_ptr) :: dcsrRowPtrB, dcsrColIndB, dcsrValB
    type(c_ptr) :: dcsrRowPtrD, dcsrColIndD, dcsrValD
    type(c_ptr) :: dcsrRowPtrC, dcsrColIndC, dcsrValC
    type(c_ptr) :: dbuffer

    ! Create hipSPARSE handle
    stat = hipsparseCreate(handle)
    if (stat /= 0) stop

    ! Create matrix descriptors
    stat = hipsparseCreateMatDescr(descrA)
    if (stat /= 0) stop
    stat = hipsparseCreateMatDescr(descrB)
    if (stat /= 0) stop
    stat = hipsparseCreateMatDescr(descrC)
    if (stat /= 0) stop
    stat = hipsparseCreateMatDescr(descrD)
    if (stat /= 0) stop

    ! Create csrgemm2 info
    stat = hipsparseCreateCsrgemm2Info(info)
    if (stat /= 0) stop

    ! Allocate device memory for A, B, and D
    stat = hipMalloc(dcsrRowPtrA, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValA, int(nnzA * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrB, int((k + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValB, int(nnzB * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrD, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrColIndD, int(nnzD * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValD, int(nnzD * 4, c_size_t))
    if (stat /= 0) stop

    stat = hipMalloc(dcsrRowPtrC, int((m + 1) * 4, c_size_t))
    if (stat /= 0) stop

    ! Copy A, B, and D to device
    stat = hipMemcpy(dcsrRowPtrA, c_loc(hcsrRowPtrA), int((m + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndA, c_loc(hcsrColIndA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValA, c_loc(hcsrValA), int(nnzA * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    stat = hipMemcpy(dcsrRowPtrB, c_loc(hcsrRowPtrB), int((k + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndB, c_loc(hcsrColIndB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValB, c_loc(hcsrValB), int(nnzB * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    stat = hipMemcpy(dcsrRowPtrD, c_loc(hcsrRowPtrD), int((m + 1) * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrColIndD, c_loc(hcsrColIndD), int(nnzD * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop
    stat = hipMemcpy(dcsrValD, c_loc(hcsrValD), int(nnzD * 4, c_size_t), hipMemcpyHostToDevice)
    if (stat /= 0) stop

    ! Get buffer size
    stat = hipsparseScsrgemm2_bufferSizeExt(handle, &
                                            m, &
                                            n, &
                                            k, &
                                            c_loc(alpha), &
                                            descrA, &
                                            nnzA, &
                                            dcsrRowPtrA, &
                                            dcsrColIndA, &
                                            descrB, &
                                            nnzB, &
                                            dcsrRowPtrB, &
                                            dcsrColIndB, &
                                            c_loc(beta), &
                                            descrD, &
                                            nnzD, &
                                            dcsrRowPtrD, &
                                            dcsrColIndD, &
                                            info, &
                                            c_loc(bufferSize))
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseScsrgemm2_bufferSizeExt failed'
        stop
    end if

    ! Allocate temporary buffer
    stat = hipMalloc(dbuffer, bufferSize)
    if (stat /= 0) stop

    ! Determine nnzC
    stat = hipsparseXcsrgemm2Nnz(handle, &
                                 m, &
                                 n, &
                                 k, &
                                 descrA, &
                                 nnzA, &
                                 dcsrRowPtrA, &
                                 dcsrColIndA, &
                                 descrB, &
                                 nnzB, &
                                 dcsrRowPtrB, &
                                 dcsrColIndB, &
                                 descrD, &
                                 nnzD, &
                                 dcsrRowPtrD, &
                                 dcsrColIndD, &
                                 descrC, &
                                 dcsrRowPtrC, &
                                 c_loc(nnzC), &
                                 info, &
                                 dbuffer)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseXcsrgemm2Nnz failed'
        stop
    end if

    ! Allocate device memory for C
    stat = hipMalloc(dcsrColIndC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop
    stat = hipMalloc(dcsrValC, int(nnzC * 4, c_size_t))
    if (stat /= 0) stop

    ! Compute C = alpha * A * B + beta * D
    stat = hipsparseScsrgemm2(handle, &
                              m, &
                              n, &
                              k, &
                              c_loc(alpha), &
                              descrA, &
                              nnzA, &
                              dcsrValA, &
                              dcsrRowPtrA, &
                              dcsrColIndA, &
                              descrB, &
                              nnzB, &
                              dcsrValB, &
                              dcsrRowPtrB, &
                              dcsrColIndB, &
                              c_loc(beta), &
                              descrD, &
                              nnzD, &
                              dcsrValD, &
                              dcsrRowPtrD, &
                              dcsrColIndD, &
                              descrC, &
                              dcsrValC, &
                              dcsrRowPtrC, &
                              dcsrColIndC, &
                              info, &
                              dbuffer)
    if (stat /= 0) then
        write(*,*) 'Error: hipsparseScsrgemm2 failed'
        stop
    end if

    ! Allocate host memory for C
    allocate(hcsrRowPtrC(m+1))
    allocate(hcsrColIndC(nnzC))
    allocate(hcsrValC(nnzC))

    ! Copy result back to host
    stat = hipMemcpy(c_loc(hcsrRowPtrC), dcsrRowPtrC, int((m + 1) * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrColIndC), dcsrColIndC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop
    stat = hipMemcpy(c_loc(hcsrValC), dcsrValC, int(nnzC * 4, c_size_t), hipMemcpyDeviceToHost)
    if (stat /= 0) stop

    ! Print result
    write(*,*) 'Matrix C (result of alpha*A*B + beta*D):'
    do i = 1, m
        start_idx = hcsrRowPtrC(i) + 1
        end_idx = hcsrRowPtrC(i + 1)
        write(*,fmt='(A,I0,A)',advance='no') 'Row ', i-1, ':'
        do stat = start_idx, end_idx
            write(*,fmt='(A,I0,A,F0.2,A)',advance='no') ' (', hcsrColIndC(stat), ',', hcsrValC(stat), ')'
        end do
        write(*,*)
    end do

    ! Clean up
    deallocate(hcsrRowPtrC)
    deallocate(hcsrColIndC)
    deallocate(hcsrValC)

    stat = hipFree(dcsrRowPtrA)
    stat = hipFree(dcsrColIndA)
    stat = hipFree(dcsrValA)
    stat = hipFree(dcsrRowPtrB)
    stat = hipFree(dcsrColIndB)
    stat = hipFree(dcsrValB)
    stat = hipFree(dcsrRowPtrD)
    stat = hipFree(dcsrColIndD)
    stat = hipFree(dcsrValD)
    stat = hipFree(dcsrRowPtrC)
    stat = hipFree(dcsrColIndC)
    stat = hipFree(dcsrValC)
    stat = hipFree(dbuffer)

    stat = hipsparseDestroyCsrgemm2Info(info)
    stat = hipsparseDestroyMatDescr(descrA)
    stat = hipsparseDestroyMatDescr(descrB)
    stat = hipsparseDestroyMatDescr(descrC)
    stat = hipsparseDestroyMatDescr(descrD)
    stat = hipsparseDestroy(handle)

end program example_hipsparse_csrgemm2

Sparse extra functions

Contents

Sparse extra functions#

hipsparseXcsrgeamNnz()#

hipsparseXcsrgeam()#

hipsparseXcsrgeam2_bufferSizeExt()#

hipsparseXcsrgeam2Nnz()#

hipsparseXcsrgeam2()#

hipsparseXcsrgemmNnz()#

hipsparseXcsrgemm()#

hipsparseXcsrgemm2_bufferSizeExt()#

hipsparseXcsrgemm2Nnz()#

hipsparseXcsrgemm2()#