Sparse Conversion Functions

This module holds all sparse conversion routines.

The sparse conversion routines describe operations on a matrix in sparse format to obtain a matrix in a different sparse format.

hipsparseXnnz()

hipsparseStatus_t hipsparseSnnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const float *A, int lda, int *nnzPerRowColumn, int *nnzTotalDevHostPtr)

hipsparseStatus_t hipsparseDnnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const double *A, int lda, int *nnzPerRowColumn, int *nnzTotalDevHostPtr)

hipsparseStatus_t hipsparseCnnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const hipComplex *A, int lda, int *nnzPerRowColumn, int *nnzTotalDevHostPtr)

hipsparseStatus_t hipsparseZnnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const hipDoubleComplex *A, int lda, int *nnzPerRowColumn, int *nnzTotalDevHostPtr)

hipsparseXnnz computes the number of nonzero elements per row or column and the total number of nonzero elements in a dense matrix.

For example, given the dense matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

then using dirA == HIPSPARSE_DIRECTION_ROW results in:

\[\begin{split} \begin{align} \text{nnzPerRowColumn} &= \begin{bmatrix} 2 & 2 & 3 \end{bmatrix} \\ \text{nnzTotalDevHostPtr} &= 7 \end{align} \end{split}\]

while using dirA == HIPSPARSE_DIRECTION_COLUMN results in:

\[\begin{split} \begin{align} \text{nnzPerRowColumn} &= \begin{bmatrix} 3 & 1 & 1 & 2 \end{bmatrix} \\ \text{nnzTotalDevHostPtr} &= 7 \end{align} \end{split}\]

The array nnzPerRowColumn must be allocated by the user before calling hipsparseXnnz and has length equal to m if dirA == HIPSPARSE_DIRECTION_ROW or n if dirA == HIPSPARSE_DIRECTION_COLUMN.

For a complete code example on its usage, see the example found with hipsparseSdense2csr().

Note

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

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXdense2csr()

hipsparseStatus_t hipsparseSdense2csr(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const float *A, int ld, const int *nnzPerRow, float *csrVal, int *csrRowPtr, int *csrColInd)

hipsparseStatus_t hipsparseDdense2csr(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const double *A, int ld, const int *nnzPerRows, double *csrCal, int *csrRowPtr, int *csrColInd)

hipsparseStatus_t hipsparseCdense2csr(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipComplex *A, int ld, const int *nnzPerRow, hipComplex *csrVal, int *csrRowPtr, int *csrColInd)

hipsparseStatus_t hipsparseZdense2csr(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipDoubleComplex *A, int ld, const int *nnzPerRows, hipDoubleComplex *csrVal, int *csrRowPtr, int *csrColInd)

hipsparseXdense2csr converts the matrix A in dense format into a sparse matrix in CSR format.

Given a dense, column ordered, matrix A with leading dimension ld where ld>=m, hipsparseXdense2csr converts the matrix to a sparse CSR format matrix. All the parameters are assumed to have been pre-allocated by the user and the arrays are filled in based on number of nonzeros per row, which can be pre-computed with hipsparseXnnz(). The desired index base in the output CSR matrix is set in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase().

As an example, if using index base zero (i.e. the default) and the dense matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

The conversion results in the CSR arrays:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 2 & 4 & 7 \end{bmatrix} \\ \text{csrColInd} &= \begin{bmatrix} 0 & 3 & 0 & 1 & 0 & 2 & 3 \end{bmatrix} \\ \text{csrVal} &= \begin{bmatrix} 1 & 2 & 3 & 4 & 5 & 6 & 7 \end{bmatrix} \\ \end{align} \end{split}\]

Example

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

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

// Dense matrix in column order
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
float hdense_A[15] = {1.0f, 0.0f, 6.0f, 2.0f, 4.0f, 0.0f, 0.0f, 5.0f, 0.0f, 3.0f, 0.0f, 7.0f, 0.0f, 0.0f, 8.0f};

int m         = 3;
int n         = 5;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;

float* ddense_A = nullptr;
hipMalloc((void**)&ddense_A, sizeof(float) * m * n);
hipMemcpy(ddense_A, hdense_A, sizeof(float) * m * n, hipMemcpyHostToDevice);

// Allocate memory for the nnz_per_row_columns array
int* dnnz_per_row;
hipMalloc((void**)&dnnz_per_row, sizeof(int) * m);

int nnz_A;
hipsparseSnnz(handle, dir, m, n, descr, ddense_A, m, dnnz_per_row, &nnz_A);

// Allocate sparse CSR matrix
int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz_A);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz_A);

hipsparseSdense2csr(handle, m, n, descr, ddense_A, m, dnnz_per_row, dcsrVal, dcsrRowPtr, dcsrColInd);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);
hipFree(dnnz_per_row);
hipFree(ddense_A);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXpruneDense2csr_bufferSize()

hipsparseStatus_t hipsparseSpruneDense2csr_bufferSize(hipsparseHandle_t handle, int m, int n, const float *A, int lda, const float *threshold, const hipsparseMatDescr_t descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneDense2csr_bufferSize(hipsparseHandle_t handle, int m, int n, const double *A, int lda, const double *threshold, const hipsparseMatDescr_t descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, size_t *pBufferSizeInBytes)

hipsparseSpruneDense2csr_bufferSize computes the the size of the user allocated temporary storage buffer used when converting a dense matrix to a pruned CSR matrix.

Specifically given an input dense column ordered matrix A, with leading dimension lda where lda>=m, the resulting pruned sparse CSR matrix C is computed using:

\[ |C(i,j)| = A(i, j) \text{ if |A(i, j)| > threshold} \]

The first step in this conversion is to determine the required user allocated buffer size using hipsparseXpruneDense2csr_bufferSize() that will be passed to the subsequent steps of the conversion. Once the buffer size has been determined the user must allocate it. This user allocated buffer is then passed to hipsparseXpruneDense2csrNnz() and hipsparseXpruneDense2csr() to complete the conversion. The user is responsible to then free the buffer once the conversion has been completed.

See hipsparseSpruneDense2csr() for a full code example.

hipsparseXpruneDense2csrNnz()

hipsparseStatus_t hipsparseSpruneDense2csrNnz(hipsparseHandle_t handle, int m, int n, const float *A, int lda, const float *threshold, const hipsparseMatDescr_t descr, int *csrRowPtr, int *nnzTotalDevHostPtr, void *buffer)

hipsparseStatus_t hipsparseDpruneDense2csrNnz(hipsparseHandle_t handle, int m, int n, const double *A, int lda, const double *threshold, const hipsparseMatDescr_t descr, int *csrRowPtr, int *nnzTotalDevHostPtr, void *buffer)

hipsparseXpruneDense2csrNnz function computes the number of nonzero elements per row and the total number of nonzero elements in a dense matrix once the elements less than the (non-negative) threshold are pruned from the matrix.

Specifically given an input dense column ordered matrix A, with leading dimension lda where lda>=m, the resulting pruned sparse CSR matrix C is computed using:

\[ |C(i,j)| = A(i, j) \text{ if |A(i, j)| > threshold} \]

First the user must determine the size of the required temporary buffer using the routine hipsparseSpruneDense2csr_bufferSize and then allocate it. Next the user allocates csrRowPtr with size m+1. Then the passes both the temporary storage buffer as well as csrRowPtr to hipsparseXpruneDense2csrNnz in order to determine the total number of non-zeros that will exist in the sparse CSR matrix C (after pruning has been performed on A) as well as fill the output CSR row pointer array csrRowPtr.

For example, given the dense matrix:

\[\begin{split} \begin{bmatrix} 6 & 2 & 3 & 7 \\ 5 & 6 & 7 & 8 \\ 5 & 4 & 8 & 1 \end{bmatrix} \end{split}\]

and the threshold value 5, the resulting matrix after pruning is:

\[\begin{split} \begin{bmatrix} 6 & 0 & 0 & 7 \\ 0 & 6 & 7 & 8 \\ 0 & 0 & 8 & 0 \end{bmatrix} \end{split}\]

and corresponding row pointer array and non-zero count:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 2 & 5 & 6 \end{bmatrix} \\ \text{nnzTotalDevHostPtr} &= 6 \end{align} \end{split}\]

The above example assumes a zero index base for the output CSR matrix. We can set the desired index base in the output CSR matrix by setting it in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase().

For a full code example on how to use this routine, see hipsparseSpruneDense2csr().

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXpruneDense2csr()

hipsparseStatus_t hipsparseSpruneDense2csr(hipsparseHandle_t handle, int m, int n, const float *A, int lda, const float *threshold, const hipsparseMatDescr_t descr, float *csrVal, const int *csrRowPtr, int *csrColInd, void *buffer)

hipsparseStatus_t hipsparseDpruneDense2csr(hipsparseHandle_t handle, int m, int n, const double *A, int lda, const double *threshold, const hipsparseMatDescr_t descr, double *csrVal, const int *csrRowPtr, int *csrColInd, void *buffer)

hipsparseXpruneDense2csr converts the matrix A in dense format into a sparse matrix in CSR format while pruning values that are less than the (non-negative) threshold. All the parameters are assumed to have been pre-allocated by the user.

Specifically given an input dense column ordered matrix A, with leading dimension lda where lda>=m, the resulting pruned sparse CSR matrix C is computed using:

\[ |C(i,j)| = A(i, j) \text{ if |A(i, j)| > threshold} \]

The user first calls hipsparseXpruneDense2csr_bufferSize to determine the size of the required user allocate temporary storage buffer. The user then allocates this buffer. Next, the user allocates csrRowPtr to have m+1 elements and then calls hipsparseXpruneDense2csrNnz() which fills in the csrRowPtr array and stores the number of elements that are larger than the pruning threshold in nnzTotalDevHostPtr. The user then allocates csrColInd and csrVal to have size nnzTotalDevHostPtr and completes the conversion by calling hipsparseXpruneDense2csr().

For example, performing these steps with the dense input matrix A:

\[\begin{split} \begin{bmatrix} 6 & 2 & 3 & 7 \\ 5 & 6 & 7 & 8 \\ 5 & 4 & 8 & 1 \end{bmatrix} \end{split}\]

and the threshold value 5, results in the pruned matrix C:

\[\begin{split} \begin{bmatrix} 6 & 0 & 0 & 7 \\ 0 & 6 & 7 & 8 \\ 0 & 0 & 8 & 0 \end{bmatrix} \end{split}\]

and corresponding CSR row, column, and values arrays:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 2 & 5 & 6 \end{bmatrix} \\ \text{csrColInd} &= \begin{bmatrix} 0 & 3 & 1 & 2 & 3 & 2 \end{bmatrix} \\ \text{csrVal} &= \begin{bmatrix} 6 & 7 & 6 & 7 & 8 & 8 \end{bmatrix} \\ \end{align} \end{split}\]

Example

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

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

// Dense matrix in column order
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
float hdense_A[15] = {1.0f, 0.0f, 6.0f, 2.0f, 4.0f, 0.0f, 0.0f, 5.0f, 0.0f, 3.0f, 0.0f, 7.0f, 0.0f, 0.0f, 8.0f};

int m         = 3;
int n         = 5;
int lda       = m;
float threshold = 4.0f;

float* ddense_A = nullptr;
hipMalloc((void**)&ddense_A, sizeof(float) * lda * n);
hipMemcpy(ddense_A, hdense_A, sizeof(float) * lda * n, hipMemcpyHostToDevice);

// Allocate sparse CSR matrix
int* dcsrRowPtr = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));

size_t bufferSize;
hipsparseSpruneDense2csr_bufferSize(handle, m, n, ddense_A, lda, &threshold, descr, nullptr, dcsrRowPtr, nullptr, &bufferSize);

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

int nnz_A;
hipsparseSpruneDense2csrNnz(handle, m, n, ddense_A, lda, &threshold, descr, dcsrRowPtr, &nnz_A, dbuffer);

int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz_A);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz_A);

hipsparseSpruneDense2csr(handle, m, n, ddense_A, lda, &threshold, descr, dcsrVal, dcsrRowPtr, dcsrColInd, dbuffer);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);
hipFree(ddense_A);
hipFree(dbuffer);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

The routine hipsparseXpruneDense2csr() is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXpruneDense2csrByPercentage_bufferSize()

hipsparseStatus_t hipsparseSpruneDense2csrByPercentage_bufferSize(hipsparseHandle_t handle, int m, int n, const float *A, int lda, float percentage, const hipsparseMatDescr_t descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, pruneInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneDense2csrByPercentage_bufferSize(hipsparseHandle_t handle, int m, int n, const double *A, int lda, double percentage, const hipsparseMatDescr_t descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, pruneInfo_t info, size_t *pBufferSizeInBytes)

hipsparseSpruneDense2csrByPercentage_bufferSize computes the size of the user allocated temporary storage buffer used when converting a dense matrix to a pruned CSR matrix where the pruning is done based on a percantage.

When converting and pruning a dense matrix A to a CSR matrix by percentage the following steps are performed. First the user calls hipsparseXpruneDense2csrByPercentage_bufferSize which determines the size of the temporary storage buffer. Once determined, this buffer must be allocated by the user. Next the user allocates the csrRowPtr array to have m+1 elements and calls hipsparseXpruneDense2csrNnzByPercentage. Finally the user finishes the conversion by allocating the csrColInd and csrVal arrays (whose size is determined by the value at nnzTotalDevHostPtr) and calling hipsparseXpruneDense2csrByPercentage.

The pruning by percentage works by first sorting the absolute values of the dense matrix A. We then determine a position in this sorted array by

\[\begin{split} pos = ceil(m*n*(percentage/100)) - 1 \\ pos = min(pos, m*n-1) \\ pos = max(pos, 0) \\ threshold = sorted_A[pos] \end{split}\]

Once we have this threshold we prune values in the dense matrix A as in hipsparseXpruneDense2csr.

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXpruneDense2csrByPercentage_bufferSizeExt()

hipsparseStatus_t hipsparseSpruneDense2csrByPercentage_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const float *A, int lda, float percentage, const hipsparseMatDescr_t descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, pruneInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneDense2csrByPercentage_bufferSizeExt(hipsparseHandle_t handle, int m, int n, const double *A, int lda, double percentage, const hipsparseMatDescr_t descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, pruneInfo_t info, size_t *pBufferSizeInBytes)

This function computes the size of the user allocated temporary storage buffer used when converting and pruning by percentage a dense matrix to a CSR matrix.

When converting and pruning a dense matrix A to a CSR matrix by percentage the following steps are performed. First the user calls hipsparseXpruneDense2csrByPercentage_bufferSizeExt which determines the size of the temporary storage buffer. Once determined, this buffer must be allocated by the user. Next the user allocates the csrRowPtr array to have m+1 elements and calls hipsparseXpruneDense2csrNnzByPercentage. Finally the user finishes the conversion by allocating the csrColInd and csrVal arrays (whos size is determined by the value at nnzTotalDevHostPtr) and calling hipsparseXpruneDense2csrByPercentage.

The pruning by percentage works by first sorting the absolute values of the dense matrix A. We then determine a position in this sorted array by

\[\begin{split} pos = ceil(m*n*(percentage/100)) - 1 \\ pos = min(pos, m*n-1) \\ pos = max(pos, 0) \\ threshold = sorted_A[pos] \end{split}\]

Once we have this threshold we prune values in the dense matrix A as in hipsparseXpruneDense2csr.

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXpruneDense2csrNnzByPercentage()

hipsparseStatus_t hipsparseSpruneDense2csrNnzByPercentage(hipsparseHandle_t handle, int m, int n, const float *A, int lda, float percentage, const hipsparseMatDescr_t descr, int *csrRowPtr, int *nnzTotalDevHostPtr, pruneInfo_t info, void *buffer)

hipsparseStatus_t hipsparseDpruneDense2csrNnzByPercentage(hipsparseHandle_t handle, int m, int n, const double *A, int lda, double percentage, const hipsparseMatDescr_t descr, int *csrRowPtr, int *nnzTotalDevHostPtr, pruneInfo_t info, void *buffer)

This function computes the number of nonzero elements per row and the total number of nonzero elements in a dense matrix when converting and pruning by percentage a dense matrix to a CSR matrix.

When converting and pruning a dense matrix A to a CSR matrix by percentage the following steps are performed. First the user calls hipsparseXpruneDense2csrByPercentage_bufferSize which determines the size of the temporary storage buffer. Once determined, this buffer must be allocated by the user. Next the user allocates the csrRowPtr array to have m+1 elements and calls hipsparseXpruneDense2csrNnzByPercentage. Finally the user finishes the conversion by allocating the csrColInd and csrVal arrays (whos size is determined by the value at nnzTotalDevHostPtr) and calling hipsparseXpruneDense2csrByPercentage.

The pruning by percentage works by first sorting the absolute values of the dense matrix A. We then determine a position in this sorted array by

\[\begin{split} pos = ceil(m*n*(percentage/100)) - 1 \\ pos = min(pos, m*n-1) \\ pos = max(pos, 0) \\ threshold = sorted_A[pos] \end{split}\]

Once we have this threshold we prune values in the dense matrix A as in hipsparseXpruneDense2csr.

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXpruneDense2csrByPercentage()

hipsparseStatus_t hipsparseSpruneDense2csrByPercentage(hipsparseHandle_t handle, int m, int n, const float *A, int lda, float percentage, const hipsparseMatDescr_t descr, float *csrVal, const int *csrRowPtr, int *csrColInd, pruneInfo_t info, void *buffer)

hipsparseStatus_t hipsparseDpruneDense2csrByPercentage(hipsparseHandle_t handle, int m, int n, const double *A, int lda, double percentage, const hipsparseMatDescr_t descr, double *csrVal, const int *csrRowPtr, int *csrColInd, pruneInfo_t info, void *buffer)

This function computes the number of nonzero elements per row and the total number of nonzero elements in a dense matrix when converting and pruning by percentage a dense matrix to a CSR matrix.

When converting and pruning a dense matrix A to a CSR matrix by percentage the following steps are performed. First the user calls hipsparseXpruneDense2csrByPercentage_bufferSize which determines the size of the temporary storage buffer. Once determined, this buffer must be allocated by the user. Next the user allocates the csrRowPtr array to have m+1 elements and calls hipsparseXpruneDense2csrNnzByPercentage. Finally the user finishes the conversion by allocating the csrColInd and csrVal arrays (whos size is determined by the value at nnzTotalDevHostPtr) and calling hipsparseXpruneDense2csrByPercentage.

The pruning by percentage works by first sorting the absolute values of the dense matrix A. We then determine a position in this sorted array by

\[\begin{split} pos = ceil(m*n*(percentage/100)) - 1 \\ pos = min(pos, m*n-1) \\ pos = max(pos, 0) \\ threshold = sorted_A[pos] \end{split}\]

Once we have this threshold we prune values in the dense matrix A as in hipsparseXpruneDense2csr.

Example

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

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

// Dense matrix in column order
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
float hdense_A[15] = {1.0f, 0.0f, 6.0f, 2.0f, 4.0f, 0.0f, 0.0f, 5.0f, 0.0f, 3.0f, 0.0f, 7.0f, 0.0f, 0.0f, 8.0f};

int m         = 3;
int n         = 5;
int lda       = m;
float percentage = 70.0f;

float* ddense_A = nullptr;
hipMalloc((void**)&ddense_A, sizeof(float) * lda * n);
hipMemcpy(ddense_A, hdense_A, sizeof(float) * lda * n, hipMemcpyHostToDevice);

// Allocate sparse CSR matrix
int* dcsrRowPtr = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));

pruneInfo_t info;
hipsparseCreatePruneInfo(&info);

size_t bufferSize;
hipsparseSpruneDense2csrByPercentage_bufferSize(handle, m, n, ddense_A, lda, percentage, descr, nullptr, dcsrRowPtr, nullptr, info, &bufferSize);

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

int nnz_A;
hipsparseSpruneDense2csrNnzByPercentage(handle, m, n, ddense_A, lda, percentage, descr, dcsrRowPtr, &nnz_A, info, dbuffer);

int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz_A);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz_A);

hipsparseSpruneDense2csrByPercentage(handle, m, n, ddense_A, lda, percentage, descr, dcsrVal, dcsrRowPtr, dcsrColInd, info, dbuffer);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);
hipFree(ddense_A);
hipFree(dbuffer);

hipsparseDestroyPruneInfo(info);
hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXdense2csc()

hipsparseStatus_t hipsparseSdense2csc(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const float *A, int ld, const int *nnzPerColumn, float *cscVal, int *cscRowInd, int *cscColPtr)

hipsparseStatus_t hipsparseDdense2csc(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const double *A, int ld, const int *nnzPerColumn, double *cscVal, int *cscRowInd, int *cscColPtr)

hipsparseStatus_t hipsparseCdense2csc(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipComplex *A, int ld, const int *nnzPerColumn, hipComplex *cscVal, int *cscRowInd, int *cscColPtr)

hipsparseStatus_t hipsparseZdense2csc(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipDoubleComplex *A, int ld, const int *nnzPerColumn, hipDoubleComplex *cscVal, int *cscRowInd, int *cscColPtr)

hipsparseXdense2csc converts the matrix A in dense format into a sparse matrix in CSC format.

Given a dense, column ordered, matrix A with leading dimension ld where ld>=m, hipsparseXdense2csc converts the matrix to a sparse CSC format matrix. All the parameters are assumed to have been pre-allocated by the user and the arrays are filled in based on number of nonzeros per row, which can be pre-computed with hipsparseXnnz(). We can set the desired index base in the output CSC matrix by setting it in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase().

As an example, if using index base zero (i.e. the default) and the dense matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

where the A values have column ordering with leading dimension ld=m:

\[\begin{split} \text{A} &= \begin{bmatrix} 1 & 3 & 5 & 0 & 4 & 0 & 0 & 0 & 6 & 2 & 0 & 7 \end{bmatrix} \\ \end{split}\]

the conversion results in the CSC arrays:

\[\begin{split} \begin{align} \text{cscRowInd} &= \begin{bmatrix} 0 & 1 & 2 & 1 & 2 & 0 & 2 \end{bmatrix} \\ \text{cscColPtr} &= \begin{bmatrix} 0 & 3 & 4 & 5 & 7 \end{bmatrix} \\ \text{cscVal} &= \begin{bmatrix} 1 & 3 & 5 & 4 & 6 & 2 & 7 \end{bmatrix} \\ \end{align} \end{split}\]

This function works very similar to hipsparseXdense2csr. See hipsparseSdense2csr() for a code example.

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXcsr2dense()

hipsparseStatus_t hipsparseScsr2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, float *A, int ld)

hipsparseStatus_t hipsparseDcsr2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, double *A, int ld)

hipsparseStatus_t hipsparseCcsr2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipComplex *csrVal, const int *csrRowPtr, const int *csrColInd, hipComplex *A, int ld)

hipsparseStatus_t hipsparseZcsr2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipDoubleComplex *csrVal, const int *csrRowPtr, const int *csrColInd, hipDoubleComplex *A, int ld)

hipsparseXcsr2dense function converts the sparse matrix in CSR format into a dense matrix.

Given the input CSR matrix of size mxn, the routine writes the matrix to the dense array A such that A has leading dimension ld and is column ordered. This means that A has size ldxn where ld>=m. All the parameters are assumed to have been pre-allocated by the user. If the input CSR matrix has index base of one, it must be set in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase() prior to calling hipsparseXcsr2dense.

For example, consider the sparse CSR matrix:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 2 & 4 & 7 \end{bmatrix} \\ \text{csrColInd} &= \begin{bmatrix} 0 & 3 & 0 & 1 & 0 & 2 & 3 \end{bmatrix} \\ \text{csrVal} &= \begin{bmatrix} 1 & 2 & 3 & 4 & 5 & 6 & 7 \end{bmatrix} \\ \end{align} \end{split}\]

hipsparseXcsr2dense is used to convert to the dense matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

where the values in the A array are column ordered:

\[\begin{split} \text{A} &= \begin{bmatrix} 1 & 3 & 5 & 0 & 4 & 0 & 0 & 0 & 6 & 2 & 0 & 7 \end{bmatrix} \\ \end{split}\]

Example

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

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

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int ld        = 3;
int nnz       = 8;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

float* ddense_A = nullptr;
hipMalloc((void**)&ddense_A, sizeof(float) * ld * n);

hipsparseScsr2dense(handle, m, n, descr, dcsrVal, dcsrRowPtr, dcsrColInd, ddense_A, ld);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);
hipFree(ddense_A);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXcsc2dense()

hipsparseStatus_t hipsparseScsc2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const float *cscVal, const int *cscRowInd, const int *cscColPtr, float *A, int ld)

hipsparseStatus_t hipsparseDcsc2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const double *cscVal, const int *cscRowInd, const int *cscColPtr, double *A, int ld)

hipsparseStatus_t hipsparseCcsc2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipComplex *cscVal, const int *cscRowInd, const int *cscColPtr, hipComplex *A, int ld)

hipsparseStatus_t hipsparseZcsc2dense(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descr, const hipDoubleComplex *cscVal, const int *cscRowInd, const int *cscColPtr, hipDoubleComplex *A, int ld)

hipsparseXcsc2dense function converts the sparse matrix in CSC format into a dense matrix.

Given the input CSC matrix of size mxn, the routine writes the matrix to the dense array A such that A has leading dimension ld and is column ordered. This means that A has size ldxn where ld>=m. All the parameters are assumed to have been pre-allocated by the user. If the input CSC matrix has index base of one, it must be set in the hipsparseMatDescr_t. See hipsparseSetMatIndexBase() prior to calling hipsparseXcsc2dense.

For example, consider the sparse CSC matrix:

\[\begin{split} \begin{align} \text{cscRowInd} &= \begin{bmatrix} 0 & 1 & 2 & 1 & 2 & 0 & 2 \end{bmatrix} \\ \text{cscColPtr} &= \begin{bmatrix} 0 & 3 & 4 & 5 & 7 \end{bmatrix} \\ \text{cscVal} &= \begin{bmatrix} 1 & 3 & 5 & 4 & 6 & 2 & 7 \end{bmatrix} \\ \end{align} \end{split}\]

hipsparseXcsc2dense is used to convert to the dense matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

where the values in the A array are column ordered:

\[\begin{split} \text{A} &= \begin{bmatrix} 1 & 3 & 5 & 0 & 4 & 0 & 0 & 0 & 6 & 2 & 0 & 7 \end{bmatrix} \\ \end{split}\]

Note

It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXcsr2bsrNnz()

hipsparseStatus_t hipsparseXcsr2bsrNnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const int *csrRowPtrA, const int *csrColIndA, int blockDim, const hipsparseMatDescr_t descrC, int *bsrRowPtrC, int *bsrNnzb)

This function computes the number of nonzero block columns per row and the total number of nonzero blocks in a sparse BSR matrix given a sparse CSR matrix as input.

Consider the matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \\ 1 & 2 & 3 & 4 \end{bmatrix} \end{split}\]

stored as a sparse CSR matrix. This function computes both the BSR row pointer array as well as the total number of non-zero blocks that results when converting the CSR matrix to the BSR format. Assuming a block dimension of 2, the above matrix once converted to BSR format looks like:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c} 1 & 0 \\ 3 & 4 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 0 \end{array} \\ \hline \begin{array}{c c} 5 & 0 \\ 1 & 2 \end{array} & \begin{array}{c c} 6 & 7 \\ 3 & 4 \end{array} \\ \end{array} \right] \end{split}\]

and the resulting BSR row pointer array and total non-zero blocks once hipsparseXcsr2bsrNnz has been called:

\[\begin{split} \begin{align} \text{bsrRowPtrC} &= \begin{bmatrix} 0 & 2 & 4 \end{bmatrix} \\ \text{bsrNnzb} &= 4 \end{align} \end{split}\]

In general, when converting a CSR matrix of size m x n to a BSR matrix, the resulting BSR matrix will have size mb x nb where mb and nb equal:

\[\begin{split} \begin{align} \text{mb} &= \text{(m - 1) / blockDim + 1} \\ \text{nb} &= \text{(n - 1) / blockDim + 1} \end{align} \end{split}\]

In particular, it may be the case that blockDim does not divide evenly into m and/or n. In these cases, the CSR matrix is expanded in size in order to fit full BSR blocks. For example, using the original CSR matrix and block dimension 3 instead of 2, the function hipsparseXcsr2bsrNnz computes the BSR row pointer array and total number of non-zero blocks for the BSR matrix:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c c} 1 & 0 & 0 \\ 3 & 4 & 0 \\ 5 & 0 & 6 \end{array} & \begin{array}{c c c} 2 & 0 & 0 \\ 0 & 0 & 0 \\ 7 & 0 & 0 \end{array} \\ \hline \begin{array}{c c c} 1 & 2 & 3 \\ 0 & 0 & 0 \\ 0 & 0 & 0 \end{array} & \begin{array}{c c c} 4 & 0 & 0 \\ 0 & 0 & 0 \\ 0 & 0 & 0 \end{array} \\ \end{array} \right] \end{split}\]

See hipsparseScsr2bsr() for full code example.

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXcsr2bsr()

hipsparseStatus_t hipsparseScsr2bsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, int blockDim, const hipsparseMatDescr_t descrC, float *bsrValC, int *bsrRowPtrC, int *bsrColIndC)

hipsparseStatus_t hipsparseDcsr2bsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, int blockDim, const hipsparseMatDescr_t descrC, double *bsrValC, int *bsrRowPtrC, int *bsrColIndC)

hipsparseStatus_t hipsparseCcsr2bsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const hipComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, int blockDim, const hipsparseMatDescr_t descrC, hipComplex *bsrValC, int *bsrRowPtrC, int *bsrColIndC)

hipsparseStatus_t hipsparseZcsr2bsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int m, int n, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrValA, const int *csrRowPtrA, const int *csrColIndA, int blockDim, const hipsparseMatDescr_t descrC, hipDoubleComplex *bsrValC, int *bsrRowPtrC, int *bsrColIndC)

Convert a sparse CSR matrix into a sparse BSR matrix.

hipsparseXcsr2bsr completes the conversion of a CSR matrix into a BSR matrix. It is assumed, that bsrValC, bsrColIndC and bsrRowPtrC are allocated. The allocation size for bsrRowPtr is computed as mb+1 where mb is the number of block rows in the BSR matrix defined as:

\[ \begin{align} \text{mb} &= \text{(m - 1) / blockDim + 1} \end{align} \]

The allocation size for bsrColIndC, i.e. bsrNnzb, is computed using hipsparseXcsr2bsrNnz() which also fills the bsrRowPtrC array. The allocation size for bsrValC is then equal to:

\[ \text{bsrNnzb * blockDim * blockDim} \]

For example, given the CSR matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \\ 1 & 2 & 3 & 4 \end{bmatrix} \end{split}\]

The resulting BSR matrix using block dimension 2 would look like:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c} 1 & 0 \\ 3 & 4 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 0 \end{array} \\ \hline \begin{array}{c c} 5 & 0 \\ 1 & 2 \end{array} & \begin{array}{c c} 6 & 7 \\ 3 & 4 \end{array} \\ \end{array} \right] \end{split}\]

The call to hipsparseXcsr2bsrNnz results in the BSR row pointer array:

\[\begin{split} \begin{align} \text{bsrRowPtrC} &= \begin{bmatrix} 0 & 2 & 4 \end{bmatrix} \\ \end{align} \end{split}\]

and the call to hipsparseXcsr2bsr completes the conversion resulting in the BSR column indices and values arrays:

\[\begin{split} \begin{align} \text{bsrColIndC} &= \begin{bmatrix} 0 & 1 & 0 & 1 \end{bmatrix} \\ \text{bsrValC} &= \begin{bmatrix} 1 & 0 & 3 & 4 & 0 & 2 & 0 & 0 & 5 & 0 & 1 & 2 & 6 & 7 & 3 & 4 \end{bmatrix} \\ \end{align} \end{split}\]

The dirA parameter determines the order of the BSR block values. The example above uses row order. Using column ordering would result instead in the BSR values array:

\[\begin{split} \text{bsrValC} &= \begin{bmatrix} 1 & 3 & 0 & 4 & 0 & 0 & 2 & 0 & 5 & 1 & 0 & 2 & 6 & 3 & 7 & 4 \end{bmatrix} \\ \end{split}\]

Example

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

// Matrix descriptor
hipsparseMatDescr_t csr_descr;
hipsparseCreateMatDescr(&csr_descr);

hipsparseMatDescr_t bsr_descr;
hipsparseCreateMatDescr(&bsr_descr);

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
int blockDim  = 3;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;

int mb = (m + blockDim - 1) / blockDim;
int nb = (n + blockDim - 1) / blockDim;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

int* dbsrRowPtr = nullptr;
hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1));

int nnzb;
hipsparseXcsr2bsrNnz(handle, dir, m, n, csr_descr, dcsrRowPtr, dcsrColInd, blockDim, bsr_descr, dbsrRowPtr, &nnzb);

int* dbsrColInd = nullptr;
float* dbsrVal = nullptr;
hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbsrVal, sizeof(float) * blockDim * blockDim * nnzb);

hipsparseScsr2bsr(handle, dir, m, n, csr_descr, dcsrVal, dcsrRowPtr, dcsrColInd, blockDim, bsr_descr, dbsrVal, dbsrRowPtr, dbsrColInd);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipFree(dbsrRowPtr);
hipFree(dbsrColInd);
hipFree(dbsrVal);

hipsparseDestroyMatDescr(csr_descr);
hipsparseDestroyMatDescr(bsr_descr);
hipsparseDestroy(handle);

Note

hipsparseXcsr2bsr requires extra temporary storage that is allocated internally if blockDim>16

hipsparseXnnz_compress()

hipsparseStatus_t hipsparseSnnz_compress(hipsparseHandle_t handle, int m, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, int *nnzPerRow, int *nnzC, float tol)

hipsparseStatus_t hipsparseDnnz_compress(hipsparseHandle_t handle, int m, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, int *nnzPerRow, int *nnzC, double tol)

hipsparseStatus_t hipsparseCnnz_compress(hipsparseHandle_t handle, int m, const hipsparseMatDescr_t descrA, const hipComplex *csrValA, const int *csrRowPtrA, int *nnzPerRow, int *nnzC, hipComplex tol)

hipsparseStatus_t hipsparseZnnz_compress(hipsparseHandle_t handle, int m, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrValA, const int *csrRowPtrA, int *nnzPerRow, int *nnzC, hipDoubleComplex tol)

This function is used as the first step in converting a CSR matrix to a compressed CSR matrix.

Given a sparse CSR matrix and a non-negative tolerance, this function computes how many entries would be left in each row of the matrix if elements less than the tolerance were removed. It also computes the total number of remaining elements in the matrix.

Specifically given an input sparse matrix A in CSR format, the resulting compressed sparse CSR matrix C is computed using:

\[ C(i,j) = A(i, j) \text{ if |A(i, j)| > tol} \]

The user first allocates nnzPerRow with size m elements. Then calling hipsparseXnnz_compress, the function fills in the nnzPerRow array and sets the total number of nonzeros found in nnzC.

See hipsparseScsr2csr_compress() for full code example.

Example

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

// Matrix descriptor
hipsparseMatDescr_t descr_A;
hipsparseCreateMatDescr(&descr_A);

//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
float tol = 4.2f;

int m     = 3;
int n     = 5;
int nnz_A = 8;

int hcsrRowPtr_A[4] = {0, 3, 5, 8};             
float hcsrVal_A[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f};

int* dcsrRowPtr_A = nullptr;
float* dcsrVal_A = nullptr;
hipMalloc((void**)&dcsrRowPtr_A, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrVal_A, sizeof(float) * nnz_A);

hipMemcpy(dcsrRowPtr_A, hcsrRowPtr_A, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrVal_A, hcsrVal_A, sizeof(float) * nnz_A, hipMemcpyHostToDevice);

// Allocate memory for the nnz_per_row array
int* dnnz_per_row;
hipMalloc((void**)&dnnz_per_row, sizeof(int) * m);

// Call snnz_compress() which fills in nnz_per_row array and finds the number
// of entries that will be in the compressed CSR matrix
int nnz_C;
hipsparseSnnz_compress(handle, m, descr_A, dcsrVal_A, dcsrRowPtr_A, dnnz_per_row, &nnz_C, tol);

hipFree(dcsrRowPtr_A);
hipFree(dcsrVal_A);
hipFree(dnnz_per_row);

hipsparseDestroyMatDescr(descr_A);
hipsparseDestroy(handle);

Note

In the case of complex matrices only the magnitude of the real part of tol is used.

hipsparseXcsr2coo()

hipsparseStatus_t hipsparseXcsr2coo(hipsparseHandle_t handle, const int *csrRowPtr, int nnz, int m, int *cooRowInd, hipsparseIndexBase_t idxBase)

Convert a sparse CSR matrix into a sparse COO matrix.

hipsparseXcsr2coo converts the CSR array containing the row offsets, that point to the start of every row, into a COO array of row indices. All arrays are assumed to be allocated by the user prior to calling hipsparseXcsr2coo.

For example, given the CSR row pointer array (assuming zero index base):

\[ \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 1 & 3 & 4 \end{bmatrix} \end{align} \]

Calling hipsparseXcsr2coo() results in the COO row indices array:

\[ \begin{align} \text{cooRowInd} &= \begin{bmatrix} 0 & 1 & 1 & 2 \end{bmatrix} \end{align} \]

Note

It can also be used to convert a CSC array containing the column offsets into a COO array of column indices.

Note

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

hipsparseXcsr2csc()

hipsparseStatus_t hipsparseScsr2csc(hipsparseHandle_t handle, int m, int n, int nnz, const float *csrSortedVal, const int *csrSortedRowPtr, const int *csrSortedColInd, float *cscSortedVal, int *cscSortedRowInd, int *cscSortedColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase)

hipsparseStatus_t hipsparseDcsr2csc(hipsparseHandle_t handle, int m, int n, int nnz, const double *csrSortedVal, const int *csrSortedRowPtr, const int *csrSortedColInd, double *cscSortedVal, int *cscSortedRowInd, int *cscSortedColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase)

hipsparseStatus_t hipsparseCcsr2csc(hipsparseHandle_t handle, int m, int n, int nnz, const hipComplex *csrSortedVal, const int *csrSortedRowPtr, const int *csrSortedColInd, hipComplex *cscSortedVal, int *cscSortedRowInd, int *cscSortedColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase)

hipsparseStatus_t hipsparseZcsr2csc(hipsparseHandle_t handle, int m, int n, int nnz, const hipDoubleComplex *csrSortedVal, const int *csrSortedRowPtr, const int *csrSortedColInd, hipDoubleComplex *cscSortedVal, int *cscSortedRowInd, int *cscSortedColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase)

Convert a sparse CSR matrix into a sparse CSC matrix.

hipsparseXcsr2csc converts a CSR matrix into a CSC matrix. hipsparseXcsr2csc can also be used to convert a CSC matrix into a CSR matrix. copyValues decides whether cscSortedVal is being filled during conversion (HIPSPARSE_ACTION_NUMERIC) or not (HIPSPARSE_ACTION_SYMBOLIC).

For example given the matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 \\ 3 & 4 & 0 & 0 \\ 5 & 0 & 6 & 7 \end{bmatrix} \end{split}\]

Represented using the sparse CSR format as:

\[ \begin{align} \text{csrSortedRowPtr} &= \begin{bmatrix} 0 & 2 & 4 & 7 \end{bmatrix} \text{csrSortedColInd} &= \begin{bmatrix} 0 & 3 & 0 & 1 & 0 & 2 & 3 \end{bmatrix} \text{csrSortedVal} &= \begin{bmatrix} 1 & 2 & 3 & 4 & 5 & 6 & 7 \end{bmatrix} \end{align} \]

this function converts to sparse CSC format:

\[ \begin{align} \text{cscSortedRowInd} &= \begin{bmatrix} 0 & 1 & 2 & 1 & 2 & 0 & 2 \end{bmatrix} \text{cscSortedColPtr} &= \begin{bmatrix} 0 & 3 & 4 & 5 & 7 \end{bmatrix} \text{cscSortedVal} &= \begin{bmatrix} 1 & 3 & 5 & 4 & 6 & 2 & 7 \end{bmatrix} \end{align} \]

The CSC arrays, cscSortedRowInd, cscSortedColPtr, and cscSortedVal must be allocated by the user prior to calling hipsparseXcsr2csc().

Example

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

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;
hipsparseAction_t action = HIPSPARSE_ACTION_NUMERIC;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

int* dcscRowInd = nullptr;
int* dcscColPtr = nullptr;
float* dcsc_val   = nullptr;
hipMalloc((void**)&dcscRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcscColPtr, sizeof(int) * (n + 1));
hipMalloc((void**)&dcsc_val, sizeof(float) * nnz);

hipsparseScsr2csc(handle, m, n, nnz, dcsrVal, dcsrRowPtr, dcsrColInd, dcsc_val, dcscRowInd, dcscColPtr, action, base);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipFree(dcscRowInd);
hipFree(dcscColPtr);
hipFree(dcsc_val);

hipsparseDestroy(handle);

Note

The resulting matrix can also be seen as the transpose of the input matrix.

Note

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

hipsparseXcsr2cscEx2_bufferSize()

hipsparseStatus_t hipsparseCsr2cscEx2_bufferSize(hipsparseHandle_t handle, int m, int n, int nnz, const void *csrVal, const int *csrRowPtr, const int *csrColInd, void *cscVal, int *cscColPtr, int *cscRowInd, hipDataType valType, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, hipsparseCsr2CscAlg_t alg, size_t *pBufferSizeInBytes)

This function computes the size of the user allocated temporary storage buffer used when converting a sparse CSR matrix into a sparse CSC matrix.

hipsparseCsr2cscEx2_bufferSize calculates the required user allocated temporary buffer needed by hipsparseCsr2cscEx2 to convert a CSR matrix into a CSC matrix. hipsparseCsr2cscEx2 can also be used to convert a CSC matrix into a CSR matrix. copyValues decides whether cscVal is being filled during conversion (HIPSPARSE_ACTION_NUMERIC) or not (HIPSPARSE_ACTION_SYMBOLIC).

Note

The resulting matrix can also be seen as the transpose of the input matrix.

Note

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

hipsparseXcsr2cscEx2()

hipsparseStatus_t hipsparseCsr2cscEx2(hipsparseHandle_t handle, int m, int n, int nnz, const void *csrVal, const int *csrRowPtr, const int *csrColInd, void *cscVal, int *cscColPtr, int *cscRowInd, hipDataType valType, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, hipsparseCsr2CscAlg_t alg, void *buffer)

Convert a sparse CSR matrix into a sparse CSC matrix.

hipsparseCsr2cscEx2 converts a CSR matrix into a CSC matrix. hipsparseCsr2cscEx2 can also be used to convert a CSC matrix into a CSR matrix. copyValues decides whether cscVal is being filled during conversion (HIPSPARSE_ACTION_NUMERIC) or not (HIPSPARSE_ACTION_SYMBOLIC).

Example

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

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;
hipsparseAction_t action = HIPSPARSE_ACTION_NUMERIC;
hipsparseCsr2CscAlg_t alg = HIPSPARSE_CSR2CSC_ALG_DEFAULT;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

int* dcscRowInd = nullptr;
int* dcscColPtr = nullptr;
float* dcsc_val   = nullptr;
hipMalloc((void**)&dcscRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcscColPtr, sizeof(int) * (n + 1));
hipMalloc((void**)&dcsc_val, sizeof(float) * nnz);

size_t bufferSize;
hipsparseCsr2cscEx2_bufferSize(handle, 
                               m, 
                               n, 
                               nnz, 
                               dcsrVal, 
                               dcsrRowPtr, 
                               dcsrColInd, 
                               dcsc_val, 
                               dcscColPtr, 
                               dcscRowInd,
                               HIP_R_32F, 
                               action, 
                               base, 
                               alg, 
                               &bufferSize);

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

hipsparseCsr2cscEx2(handle, 
                    m, 
                    n, 
                    nnz, 
                    dcsrVal, 
                    dcsrRowPtr, 
                    dcsrColInd, 
                    dcsc_val, 
                    dcscColPtr, 
                    dcscRowInd, 
                    HIP_R_32F, 
                    action, 
                    base, 
                    alg, 
                    dbuffer);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipFree(dcscRowInd);
hipFree(dcscColPtr);
hipFree(dcsc_val);

hipFree(dbuffer);

hipsparseDestroy(handle);

Note

The resulting matrix can also be seen as the transpose of the input matrix.

Note

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

hipsparseXcsr2hyb()

hipsparseStatus_t hipsparseScsr2hyb(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const float *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, hipsparseHybMat_t hybA, int userEllWidth, hipsparseHybPartition_t partitionType)

hipsparseStatus_t hipsparseDcsr2hyb(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const double *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, hipsparseHybMat_t hybA, int userEllWidth, hipsparseHybPartition_t partitionType)

hipsparseStatus_t hipsparseCcsr2hyb(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const hipComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, hipsparseHybMat_t hybA, int userEllWidth, hipsparseHybPartition_t partitionType)

hipsparseStatus_t hipsparseZcsr2hyb(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrSortedValA, const int *csrSortedRowPtrA, const int *csrSortedColIndA, hipsparseHybMat_t hybA, int userEllWidth, hipsparseHybPartition_t partitionType)

Convert a sparse CSR matrix into a sparse HYB matrix.

hipsparseXcsr2hyb converts a CSR matrix into a HYB matrix. It is assumed that hyb has been initialized with hipsparseCreateHybMat().

Example

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

hipsparseMatDescr_t descr;
hipsparseCreateMatDescr(&descr);

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
int userEllWidth = 2;
hipsparseHybPartition_t partitionType = HIPSPARSE_HYB_PARTITION_AUTO;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

hipsparseHybMat_t hyb;
hipsparseCreateHybMat(&hyb);

hipsparseScsr2hyb(handle, m, n, descr, dcsrVal, dcsrRowPtr, dcsrColInd, hyb, userEllWidth, partitionType);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipsparseDestroyHybMat(hyb);    
hipsparseDestroyMatDescr(descr);

hipsparseDestroy(handle);

Note

This function requires a significant amount of storage for the HYB matrix, depending on the matrix structure.

Note

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

hipsparseXgebsr2gebsc_bufferSize()

hipsparseStatus_t hipsparseSgebsr2gebsc_bufferSize(hipsparseHandle_t handle, int mb, int nb, int nnzb, const float *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDgebsr2gebsc_bufferSize(hipsparseHandle_t handle, int mb, int nb, int nnzb, const double *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseCgebsr2gebsc_bufferSize(hipsparseHandle_t handle, int mb, int nb, int nnzb, const hipComplex *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseZgebsr2gebsc_bufferSize(hipsparseHandle_t handle, int mb, int nb, int nnzb, const hipDoubleComplex *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

Convert a sparse GEBSR matrix into a sparse GEBSC matrix.

hipsparseXgebsr2gebsc_bufferSize returns the size of the temporary storage buffer required by hipsparseXgebsr2gebsc() and is the first step in converting a sparse matrix in GEBSR format to a sparse matrix in GEBSC format. Once the size of the temporary storage buffer has been determined, it must be allocated by the user.

See hipsparseSgebsr2gebsc() for a complete code example.

hipsparseXgebsr2gebsc()

hipsparseStatus_t hipsparseSgebsr2gebsc(hipsparseHandle_t handle, int mb, int nb, int nnzb, const float *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, float *bscVal, int *bscRowInd, int *bscColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, void *temp_buffer)

hipsparseStatus_t hipsparseDgebsr2gebsc(hipsparseHandle_t handle, int mb, int nb, int nnzb, const double *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, double *bscVal, int *bscRowInd, int *bscColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, void *temp_buffer)

hipsparseStatus_t hipsparseCgebsr2gebsc(hipsparseHandle_t handle, int mb, int nb, int nnzb, const hipComplex *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, hipComplex *bscVal, int *bscRowInd, int *bscColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, void *temp_buffer)

hipsparseStatus_t hipsparseZgebsr2gebsc(hipsparseHandle_t handle, int mb, int nb, int nnzb, const hipDoubleComplex *bsrVal, const int *bsrRowPtr, const int *bsrColInd, int rowBlockDim, int colBlockDim, hipDoubleComplex *bscVal, int *bscRowInd, int *bscColPtr, hipsparseAction_t copyValues, hipsparseIndexBase_t idxBase, void *temp_buffer)

Convert a sparse GEBSR matrix into a sparse GEBSC matrix.

hipsparseXgebsr2gebsc converts a GEBSR matrix into a GEBSC matrix. hipsparseXgebsr2gebsc can also be used to convert a GEBSC matrix into a GEBSR matrix. copyValues decides whether bscVal is being filled during conversion (HIPSPARSE_ACTION_NUMERIC) or not (HIPSPARSE_ACTION_SYMBOLIC).

hipsparseXgebsr2gebsc requires extra temporary storage buffer that has to be allocated by the user. Storage buffer size can be determined by hipsparseXgebsr2gebsc_bufferSize().

For example, given the GEBSR matrix:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c} 1 & 2 \\ 3 & 4 \\ 6 & 0 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 0 \\ 3 & 4 \end{array} \\ \hline \begin{array}{c c} 5 & 0 \\ 1 & 2 \\ 3 & 4 \end{array} & \begin{array}{c c} 6 & 7 \\ 3 & 4 \\ 3 & 4 \end{array} \\ \end{array} \right] \end{split}\]

represented with the arrays:

\[\begin{split} \begin{align} \text{bsrRowPtr} &= \begin{bmatrix} 0 & 2 & 4 \end{bmatrix} \\ \text{bsrColInd} &= \begin{bmatrix} 0 & 1 & 0 & 1 \end{bmatrix} \\ \text{bsrVal} &= \begin{bmatrix} 1 & 2 & 3 & 4 & 6 & 0 & 0 & 2 & 0 & 0 & 3 & 4 & 5 & 0 & 1 & 2 & 3 & 4 & 6 & 7 & 3 & 4 & 3 & 4 \end{bmatrix} \end{align} \end{split}\]

this function converts the matrix to GEBSC format:

\[\begin{split} \begin{align} \text{bscRowInd} &= \begin{bmatrix} 0 & 1 & 0 & 1 \end{bmatrix} \\ \text{bscColPtr} &= \begin{bmatrix} 0 & 2 & 4 \end{bmatrix} \\ \text{bscVal} &= \begin{bmatrix} 1 & 2 & 3 & 4 & 6 & 0 & 5 & 0 & 1 & 2 & 3 & 4 & 0 & 2 & 0 & 0 & 3 & 4 & 6 & 7 & 3 & 4 & 3 & 4 \end{bmatrix} \end{align} \end{split}\]

The GEBSC arrays, bscRowInd, bscColPtr, and bscVal must be allocated by the user prior to calling hipsparseXgebsr2gebsc(). The bscRowInd array has size nnzb, the bscColPtr array has size nb+1, and the bscVal array has size nnzb*rowBlockDim*colBlockDim.

Example

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

// Sparse matrix in BSR format
//     1 2 | 0 3 | 0 0 
//     0 4 | 5 0 | 0 1
// A = 6 0 | 0 7 | 8 0
//     --------------- 
//     0 0 | 3 0 | 2 2
//     1 0 | 0 0 | 4 3 
//     7 2 | 0 0 | 1 4
int hbsrRowPtr[3] = {0, 3, 6};
int hbsrColInd[6] = {0, 1, 2, 0, 1, 2};
float hbsrVal[36]  = {1.0f, 2.0f, 0.0f, 4.0f, 6.0f, 0.0f, 
                       0.0f, 3.0f, 5.0f, 0.0f, 0.0f, 7.0f, 
                       0.0f, 0.0f, 0.0f, 1.0f, 8.0f, 0.0f, 
                       0.0f, 0.0f, 1.0f, 0.0f, 7.0f, 2.0f,
                       3.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f,
                       2.0f, 2.0f, 4.0f, 3.0f, 1.0f, 4.0f}; 

int m           = 6;
int n           = 6;
int rowBlockDim = 3;
int colBlockDim = 2;
int nnzb        = 6;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;
hipsparseAction_t    action = HIPSPARSE_ACTION_NUMERIC;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;

int mb = (m + rowBlockDim - 1) / rowBlockDim;
int nb = (n + colBlockDim - 1) / colBlockDim;

int* dbsrRowPtr = nullptr;
int* dbsrColInd = nullptr;
float* dbsrVal = nullptr;
hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1));
hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbsrVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb);

hipMemcpy(dbsrRowPtr, hbsrRowPtr, sizeof(int) * (mb + 1), hipMemcpyHostToDevice);
hipMemcpy(dbsrColInd, hbsrColInd, sizeof(int) * nnzb, hipMemcpyHostToDevice);
hipMemcpy(dbsrVal, hbsrVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb, hipMemcpyHostToDevice);

int* dbscRowInd = nullptr;
int* dbscColPtr = nullptr;
float* dbscVal = nullptr;
hipMalloc((void**)&dbscRowInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbscColPtr, sizeof(int) * (nb + 1));
hipMalloc((void**)&dbscVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb);

size_t bufferSize;
hipsparseSgebsr2gebsc_bufferSize(handle, 
                                 mb, 
                                 nb, 
                                 nnzb, 
                                 dbsrVal, 
                                 dbsrRowPtr, 
                                 dbsrColInd, 
                                 rowBlockDim, 
                                 colBlockDim, 
                                 &bufferSize);

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

hipsparseSgebsr2gebsc(handle, 
                      mb, 
                      nb, 
                      nnzb, 
                      dbsrVal, 
                      dbsrRowPtr, 
                      dbsrColInd, 
                      rowBlockDim, 
                      colBlockDim, 
                      dbscVal, 
                      dbscRowInd, 
                      dbscColPtr, 
                      action, 
                      base, 
                      dbuffer);

hipFree(dbsrRowPtr);
hipFree(dbsrColInd);
hipFree(dbsrVal);

hipFree(dbscRowInd);
hipFree(dbscColPtr);
hipFree(dbscVal);

hipFree(dbuffer);

hipsparseDestroy(handle);

Note

The resulting matrix can also be seen as the transpose of the input matrix.

Note

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

hipsparseXcsr2gebsr_bufferSize()

hipsparseStatus_t hipsparseScsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDcsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseCcsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const hipComplex *csrVal, const int *csrRowPtr, const int *csrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseZcsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const hipDoubleComplex *csrVal, const int *csrRowPtr, const int *csrColInd, int rowBlockDim, int colBlockDim, size_t *pBufferSizeInBytes)

Convert a sparse CSR matrix into a sparse GEBSR matrix.

hipsparseXcsr2gebsr_bufferSize returns the size of the temporary buffer that is required by hipsparseXcsr2gebcsrNnz and hipsparseXcsr2gebcsr. Once the temporary buffer size has been determined, it must be allocated by the user prior to calling hipsparseXcsr2gebcsrNnz and hipsparseXcsr2gebcsr.

See hipsparseScsr2gebsr() for complete code example.

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXcsr2gebsrNnz()

hipsparseStatus_t hipsparseXcsr2gebsrNnz(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const int *csrRowPtr, const int *csrColInd, const hipsparseMatDescr_t bsr_descr, int *bsrRowPtr, int rowBlockDim, int colBlockDim, int *bsrNnzDevhost, void *pbuffer)

This function computes the number of nonzero block columns per row and the total number of nonzero blocks in a sparse GEBSR matrix given a sparse CSR matrix as input.

This is the second step in conveting a CSR matrix to a GEBSR matrix. The user must first call hipsparseXcsr2gebsr_bufferSize to determine the size of the required temporary storage buffer. The user then allocates this buffer as well as the bsrRowPtr array ( size mb+1 ) and passes both to hipsparseXcsr2gebsrNnz(). This second step then computes the number of nonzero block columns per row and the total number of nonzero blocks.

In general, when converting a CSR matrix of size m x n to a GEBSR matrix, the resulting GEBSR matrix will have size mb x nb where mb and nb equal:

\[\begin{split} \begin{align} \text{mb} &= \text{(m - 1) / rowBlockDim + 1} \\ \text{nb} &= \text{(n - 1) / colBlockDim + 1} \end{align} \end{split}\]

For example given a matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 & 4 & 0 \\ 3 & 4 & 0 & 0 & 5 & 1 \\ 5 & 0 & 6 & 7 & 6 & 2 \end{bmatrix} \end{split}\]

represented in CSR format with the arrays:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 3 & 7 & 12 \end{bmatrix} \\ \text{csrColInd} &= \begin{bmatrix} 0 & 3 & 4 & 0 & 1 & 4 & 5 & 0 & 2 & 3 & 4 & 5 \end{bmatrix} \\ \text{csrVal} &= \begin{bmatrix} 1 & 2 & 4 & 3 & 4 & 5 & 1 & 5 & 6 & 7 & 6 & 2 \end{bmatrix} \end{align} \end{split}\]

the bsrRowPtr array and total nonzero block count will be filled with:

\[\begin{split} \begin{align} \text{bsrRowPtr} &= \begin{bmatrix} 0 & 3 \end{bmatrix} \\ \text{*bsrNnzDevhost} &= 3 \end{align} \end{split}\]

after calling hipsparseXcsr2gebsrNnz with rowBlockDim=3 and colBlockDim=2.

It may be the case that rowBlockDim does not divide evenly into m and/or that colBlockDim does not divide evenly into n. In these cases, the CSR matrix is expanded in size in order to fit full GEBSR blocks. For example, using the original CSR matrix but this time with rowBlockDim=2 and colBlockDim=3, the function hipsparseXcsr2gebsrNnz computes the GEBSR row pointer array and total number of non-zero blocks for the GEBSR matrix:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c c} 1 & 0 & 0 \\ 3 & 4 & 0 \end{array} & \begin{array}{c c c} 2 & 4 & 0 \\ 0 & 5 & 1 \end{array} \\ \hline \begin{array}{c c c} 5 & 0 & 6 \\ 0 & 0 & 0 \end{array} & \begin{array}{c c c} 7 & 6 & 2 \\ 0 & 0 & 0 \end{array} \end{array} \right] \end{split}\]

See hipsparseScsr2gebsr() for full code example.

Note

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

hipsparseXcsr2gebsr()

hipsparseStatus_t hipsparseScsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const float *csrVal, const int *csrRowPtr, const int *csrColInd, const hipsparseMatDescr_t bsr_descr, float *bsrVal, int *bsrRowPtr, int *bsrColInd, int rowBlockDim, int colBlockDim, void *pbuffer)

hipsparseStatus_t hipsparseDcsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const double *csrVal, const int *csrRowPtr, const int *csrColInd, const hipsparseMatDescr_t bsr_descr, double *bsrVal, int *bsrRowPtr, int *bsrColInd, int rowBlockDim, int colBlockDim, void *pbuffer)

hipsparseStatus_t hipsparseCcsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const hipComplex *csrVal, const int *csrRowPtr, const int *csrColInd, const hipsparseMatDescr_t bsr_descr, hipComplex *bsrVal, int *bsrRowPtr, int *bsrColInd, int rowBlockDim, int colBlockDim, void *pbuffer)

hipsparseStatus_t hipsparseZcsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dir, int m, int n, const hipsparseMatDescr_t csr_descr, const hipDoubleComplex *csrVal, const int *csrRowPtr, const int *csrColInd, const hipsparseMatDescr_t bsr_descr, hipDoubleComplex *bsrVal, int *bsrRowPtr, int *bsrColInd, int rowBlockDim, int colBlockDim, void *pbuffer)

Convert a sparse CSR matrix into a sparse GEBSR matrix.

hipsparseXcsr2gebsr converts a CSR matrix into a GEBSR matrix. It is assumed, that bsrVal, bsrColInd and bsrRowPtr are allocated. Allocation size for bsrRowPtr is computed as mb+1 where mb is the number of block rows in the GEBSR matrix. The number of nonzero blocks in the resulting GEBSR matrix is computed using hipsparseXcsr2gebsrNnz which also fills in bsrRowPtr.

In more detail, hipsparseXcsr2gebsr is the third and final step on the conversion from CSR to GEBSR. The user first determines the size of the required user allocated temporary storage buffer using hipsparseXcsr2gebsr_bufferSize. The user then allocates this buffer as well as the row pointer array bsrRowPtr with size mb+1, where mb is the number of block rows in the GEBSR matrix and nb is the number of block columns in GEBSR matrix:

\[\begin{split} \begin{align} \text{mb} &= \text{(m - 1) / rowBlockDim + 1} \\ \text{nb} &= \text{(n - 1) / colBlockDim + 1} \end{align} \end{split}\]

Both the temporary storage buffer and the GEBSR row pointer array are then passed to hipsparseXcsr2gebsrNnz which fills the GEBSR row pointer array bsrRowPtr and also computes the number of nonzero blocks, bsr_nnz, that will exist in the GEBSR matrix. The user then allocates both the GEBSR column indices array bsrColInd with size bsr_nnz as well as the GEBSR values array bsrVal with size bsr_nnz*rowBlockDim*colBlockDim. Finally, with all arrays allocated, the conversion is completed by calling hipsparseXcsr2gebsr.

For example, assuming the matrix:

\[\begin{split} \begin{bmatrix} 1 & 0 & 0 & 2 & 4 & 0 \\ 3 & 4 & 0 & 0 & 5 & 1 \\ 5 & 0 & 6 & 7 & 6 & 2 \end{bmatrix} \end{split}\]

represented in CSR format with the arrays:

\[\begin{split} \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 3 & 7 & 12 \end{bmatrix} \\ \text{csrColInd} &= \begin{bmatrix} 0 & 3 & 4 & 0 & 1 & 4 & 5 & 0 & 2 & 3 & 4 & 5 \end{bmatrix} \\ \text{csrVal} &= \begin{bmatrix} 1 & 2 & 4 & 3 & 4 & 5 & 1 & 5 & 6 & 7 & 6 & 2 \end{bmatrix} \end{align} \end{split}\]

then using rowBlockDim=3 and colBlockDim=2, the final GEBSR matrix is:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c} 1 & 0 \\ 3 & 4 \\ 3 & 0 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 0 \\ 6 & 7 \end{array} & \begin{array}{c c} 4 & 0 \\ 5 & 1 \\ 6 & 2 \end{array} \end{array} \right] \end{split}\]

and is represented with the arrays:

\[\begin{split} \begin{align} \text{bsrRowPtr} &= \begin{bmatrix} 0 & 3 \end{bmatrix} \\ \text{bsrColInd} &= \begin{bmatrix} 0 & 1 & 2 \end{bmatrix} \\ \text{bsrVal} &= \begin{bmatrix} 1 & 0 & 3 & 4 & 3 & 0 & 0 & 2 & 0 & 0 & 6 & 7 & 4 & 0 & 5 & 1 & 6 & 2 \end{bmatrix} \end{align} \end{split}\]

The above example assumes that the blocks are row ordered. If instead the blocks are column ordered, the bsrVal arrays becomes:

\[ \begin{align} \text{bsrVal} &= \begin{bmatrix} 1 & 3 & 3 & 0 & 4 & 0 & 0 & 0 & 6 & 2 & 0 & 7 & 4 & 5 & 6 & 0 & 1 & 2 \end{bmatrix} \end{align} \]

The block order direction is determined by dir.

It may be the case that rowBlockDim does not divide evenly into m and/or that colBlockDim does not divide evenly into n. In these cases, the CSR matrix is expanded in size in order to fit full GEBSR blocks. For example, using the original CSR matrix but this time with rowBlockDim=2 and colBlockDim=3, the resulting GEBSR matrix would looks like:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c c} 1 & 0 & 0 \\ 3 & 4 & 0 \end{array} & \begin{array}{c c c} 2 & 4 & 0 \\ 0 & 5 & 1 \end{array} \\ \hline \begin{array}{c c c} 5 & 0 & 6 \\ 0 & 0 & 0 \end{array} & \begin{array}{c c c} 7 & 6 & 2 \\ 0 & 0 & 0 \end{array} \end{array} \right] \end{split}\]

Example

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

hipsparseMatDescr_t csr_descr;
hipsparseCreateMatDescr(&csr_descr);

hipsparseMatDescr_t bsr_descr;
hipsparseCreateMatDescr(&bsr_descr);

// Sparse matrix in CSR format
//     1 2 0 3 0 0 
//     0 4 5 0 0 1
// A = 6 0 0 7 8 0
//     0 0 3 0 2 2
//     1 0 0 0 4 3 
//     7 2 0 0 1 4
int hcsrRowPtr[7] = {0, 3, 6, 9, 12, 15, 19};
int hcsrColInd[19] = {0, 1, 3, 1, 2, 5, 0, 3, 4, 2, 4, 5, 0, 4, 5, 0, 1, 4, 5};
float hcsrVal[19]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 1.0f, 6.0f, 7.0f, 8.0f, 3.0f, 2.0f, 2.0f,
                       1.0f, 4.0f, 3.0f, 7.0f, 2.0f, 1.0f, 4.0f}; 

int m           = 6;
int n           = 6;
int nnz         = 19;
int rowBlockDim = 3;
int colBlockDim = 2;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;

int mb = (m + rowBlockDim - 1) / rowBlockDim;
int nb = (n + colBlockDim - 1) / colBlockDim;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

int* dbsrRowPtr = nullptr;
hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1));

size_t bufferSize;
hipsparseScsr2gebsr_bufferSize(handle, 
                               dir, 
                               m, 
                               n, 
                               csr_descr, 
                               dcsrVal, 
                               dcsrRowPtr, 
                               dcsrColInd, 
                               rowBlockDim, 
                               colBlockDim, 
                               &bufferSize);

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

int nnzb;
hipsparseXcsr2gebsrNnz(handle, 
                       dir, 
                       m, 
                       n, 
                       csr_descr, 
                       dcsrRowPtr, 
                       dcsrColInd, 
                       bsr_descr, 
                       dbsrRowPtr, 
                       rowBlockDim, 
                       colBlockDim, 
                       &nnzb, 
                       dbuffer);

int* dbsrColInd = nullptr;
float* dbsrVal = nullptr;
hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbsrVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb);

hipsparseScsr2gebsr(handle, 
                    dir, 
                    m, 
                    n, 
                    csr_descr, 
                    dcsrVal, 
                    dcsrRowPtr, 
                    dcsrColInd, 
                    bsr_descr, 
                    dbsrVal, 
                    dbsrRowPtr, 
                    dbsrColInd, 
                    rowBlockDim, 
                    colBlockDim, 
                    dbuffer);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipFree(dbsrRowPtr);
hipFree(dbsrColInd);
hipFree(dbsrVal);

hipFree(dbuffer);

hipsparseDestroyMatDescr(csr_descr);
hipsparseDestroyMatDescr(bsr_descr);
hipsparseDestroy(handle);

hipsparseXbsr2csr()

hipsparseStatus_t hipsparseSbsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const float *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int blockDim, const hipsparseMatDescr_t descrC, float *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseDbsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const double *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int blockDim, const hipsparseMatDescr_t descrC, double *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseCbsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const hipComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int blockDim, const hipsparseMatDescr_t descrC, hipComplex *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseZbsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int blockDim, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrValC, int *csrRowPtrC, int *csrColIndC)

Convert a sparse BSR matrix into a sparse CSR matrix.

hipsparseXbsr2csr converts a BSR matrix into a CSR matrix. It is assumed, that csrValC, csrColIndC and csrRowPtrC are allocated. Allocation size for csrRowPtrC is computed by the number of block rows multiplied by the block dimension plus one. Allocation for csrValC and csrColInd is computed by the the number of blocks in the BSR matrix multiplied by the block dimension squared.

For example, given the BSR matrix using block dimension 2:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c} 1 & 0 \\ 3 & 4 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 0 \end{array} \\ \hline \begin{array}{c c} 5 & 0 \\ 1 & 2 \end{array} & \begin{array}{c c} 6 & 7 \\ 3 & 4 \end{array} \\ \end{array} \right] \end{split}\]

The resulting CSR matrix row pointer, column indices, and values arrays are:

\[\begin{split} \begin{align} \text{csrRowPtrC} &= \begin{bmatrix} 0 & 4 & 8 & 12 & 16 \end{bmatrix} \\ \text{csrColIndC} &= \begin{bmatrix} 0 & 1 & 2 & 3 & 0 & 1 & 2 & 3 & 0 & 1 & 2 & 3 & 0 & 1 & 2 & 3 \end{bmatrix} \\ \text{csrValC} &= \begin{bmatrix} 1 & 0 & 0 & 2 & 3 & 4 & 0 & 0 & 5 & 0 & 6 & 7 & 1 & 2 & 3 & 4 \end{bmatrix} \\ \end{align} \end{split}\]

Example

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

hipsparseMatDescr_t csr_descr;
hipsparseCreateMatDescr(&csr_descr);

hipsparseMatDescr_t bsr_descr;
hipsparseCreateMatDescr(&bsr_descr);

// Sparse matrix in BSR format
//     1 2 | 0 3 | 0 0 
//     0 4 | 5 0 | 0 1
//     ---------------
// A = 6 0 | 0 7 | 8 0 
//     0 0 | 3 0 | 2 2
//     ---------------
//     1 0 | 0 0 | 4 3 
//     7 2 | 0 0 | 1 4
int hbsrRowPtr[4] = {0, 3, 6, 8};
int hbsrColInd[8] = {0, 1, 2, 0, 1, 2, 0, 2};
float hbsrVal[32]  = {1.0f, 2.0f, 0.0f, 4.0f, 
                        0.0f, 3.0f, 5.0f, 0.0f, 
                        0.0f, 0.0f, 0.0f, 1.0f,
                        6.0f, 0.0f, 0.0f, 0.0f, 
                        0.0f, 7.0f, 3.0f, 0.0f,
                        8.0f, 0.0f, 2.0f, 2.0f,
                        1.0f, 0.0f, 7.0f, 2.0f,
                        4.0f, 3.0f, 1.0f, 4.0f}; 

int m        = 6;
int n        = 6;
int nnz      = 32;
int mb       = 3;
int nb       = 3;
int nnzb     = 8;
int blockDim = 2;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;

int* dbsrRowPtr = nullptr;
int* dbsrColInd = nullptr;
float* dbsrVal = nullptr;
hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1));
hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbsrVal, sizeof(float) * blockDim * blockDim * nnzb);

hipMemcpy(dbsrRowPtr, hbsrRowPtr, sizeof(int) * (mb + 1), hipMemcpyHostToDevice);
hipMemcpy(dbsrColInd, hbsrColInd, sizeof(int) * nnzb, hipMemcpyHostToDevice);
hipMemcpy(dbsrVal, hbsrVal, sizeof(float) * blockDim * blockDim * nnzb, hipMemcpyHostToDevice);

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipsparseSbsr2csr(handle, 
                  dir, 
                  mb, 
                  nb, 
                  bsr_descr, 
                  dbsrVal, 
                  dbsrRowPtr, 
                  dbsrColInd, 
                  blockDim, 
                  csr_descr, 
                  dcsrVal, 
                  dcsrRowPtr, 
                  dcsrColInd);

hipFree(dbsrRowPtr);
hipFree(dbsrColInd);
hipFree(dbsrVal);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipsparseDestroyMatDescr(csr_descr);
hipsparseDestroyMatDescr(bsr_descr);
hipsparseDestroy(handle);

Note

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

hipsparseXgebsr2csr()

hipsparseStatus_t hipsparseSgebsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const float *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDim, int colBlockDim, const hipsparseMatDescr_t descrC, float *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseDgebsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const double *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDim, int colBlockDim, const hipsparseMatDescr_t descrC, double *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseCgebsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const hipComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDim, int colBlockDim, const hipsparseMatDescr_t descrC, hipComplex *csrValC, int *csrRowPtrC, int *csrColIndC)

hipsparseStatus_t hipsparseZgebsr2csr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDim, int colBlockDim, const hipsparseMatDescr_t descrC, hipDoubleComplex *csrValC, int *csrRowPtrC, int *csrColIndC)

Convert a sparse GEBSR matrix into a sparse CSR matrix.

hipsparseXgebsr2csr converts a GEBSR matrix into a CSR matrix. It is assumed, that csrValC, csrColIndC and csrRowPtrC are already allocated prior to calling hipsparseXgebsr2csr. Allocation size for csrRowPtrC equals m+1 where:

\[\begin{split} \begin{align} \text{m} &= \text{mb * rowBlockDim} \\ \text{n} &= \text{nb * colBlockDim} \end{align} \end{split}\]

Allocation size for csrValC and csrColIndC is computed by the the number of blocks in the GEBSR matrix, nnzb, multiplied by the product of the block dimensions, i.e. nnz=nnzb*rocBlockDim*colBlockDim.

For example, given the GEBSR matrix:

\[\begin{split} \left[ \begin{array}{c | c | c} \begin{array}{c c} 6 & 2 \\ 1 & 4 \\ 5 & 4 \end{array} & \begin{array}{c c} 0 & 3 \\ 5 & 0 \\ 0 & 7 \end{array} & \begin{array}{c c} 0 & 0 \\ 0 & 0 \\ 0 & 0 \end{array} \\ \hline \begin{array}{c c} 0 & 0 \\ 0 & 0 \\ 0 & 0 \end{array} & \begin{array}{c c} 3 & 0 \\ 0 & 0 \\ 0 & 7 \end{array} & \begin{array}{c c} 2 & 2 \\ 4 & 3 \\ 1 & 4 \end{array} \\ \end{array} \right] \end{split}\]

Example

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

hipsparseMatDescr_t csr_descr;
hipsparseCreateMatDescr(&csr_descr);

hipsparseMatDescr_t bsr_descr;
hipsparseCreateMatDescr(&bsr_descr);

// Sparse matrix in GEBSR format
//     1 2 | 0 3 | 0 0 
//     0 4 | 5 0 | 0 1
// A = 6 0 | 0 7 | 8 0
//     --------------- 
//     0 0 | 3 0 | 2 2
//     1 0 | 0 0 | 4 3 
//     7 2 | 0 0 | 1 4
int hbsrRowPtr[3] = {0, 3, 6};
int hbsrColInd[6] = {0, 1, 2, 0, 1, 2};
float hbsrVal[36]  = {1.0f, 2.0f, 0.0f, 4.0f, 6.0f, 0.0f, 
                       0.0f, 3.0f, 5.0f, 0.0f, 0.0f, 7.0f,
                       0.0f, 0.0f, 0.0f, 1.0f, 8.0f, 0.0f, 
                       0.0f, 0.0f, 1.0f, 0.0f, 7.0f, 2.0f,
                       3.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f, 
                       2.0f, 2.0f, 4.0f, 3.0f, 1.0f, 4.0f}; 

int m        = 6;
int n        = 6;
int nnz      = 36;
int mb       = 2;
int nb       = 3;
int nnzb     = 6;
int rowBlockDim = 3;
int colBlockDim = 2;
hipsparseDirection_t dir = HIPSPARSE_DIRECTION_ROW;

int* dbsrRowPtr = nullptr;
int* dbsrColInd = nullptr;
float* dbsrVal = nullptr;
hipMalloc((void**)&dbsrRowPtr, sizeof(int) * (mb + 1));
hipMalloc((void**)&dbsrColInd, sizeof(int) * nnzb);
hipMalloc((void**)&dbsrVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb);

hipMemcpy(dbsrRowPtr, hbsrRowPtr, sizeof(int) * (mb + 1), hipMemcpyHostToDevice);
hipMemcpy(dbsrColInd, hbsrColInd, sizeof(int) * nnzb, hipMemcpyHostToDevice);
hipMemcpy(dbsrVal, hbsrVal, sizeof(float) * rowBlockDim * colBlockDim * nnzb, hipMemcpyHostToDevice);

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipsparseSgebsr2csr(handle, 
                    dir, 
                    mb, 
                    nb, 
                    bsr_descr, 
                    dbsrVal, 
                    dbsrRowPtr, 
                    dbsrColInd, 
                    rowBlockDim, 
                    colBlockDim, 
                    csr_descr, 
                    dcsrVal, 
                    dcsrRowPtr, 
                    dcsrColInd);

hipFree(dbsrRowPtr);
hipFree(dbsrColInd);
hipFree(dbsrVal);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);

hipsparseDestroyMatDescr(csr_descr);
hipsparseDestroyMatDescr(bsr_descr);
hipsparseDestroy(handle);

Note

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

hipsparseXcsr2csr_compress()

hipsparseStatus_t hipsparseScsr2csr_compress(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrColIndA, const int *csrRowPtrA, int nnzA, const int *nnzPerRow, float *csrValC, int *csrColIndC, int *csrRowPtrC, float tol)

hipsparseStatus_t hipsparseDcsr2csr_compress(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrColIndA, const int *csrRowPtrA, int nnzA, const int *nnzPerRow, double *csrValC, int *csrColIndC, int *csrRowPtrC, double tol)

hipsparseStatus_t hipsparseCcsr2csr_compress(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const hipComplex *csrValA, const int *csrColIndA, const int *csrRowPtrA, int nnzA, const int *nnzPerRow, hipComplex *csrValC, int *csrColIndC, int *csrRowPtrC, hipComplex tol)

hipsparseStatus_t hipsparseZcsr2csr_compress(hipsparseHandle_t handle, int m, int n, const hipsparseMatDescr_t descrA, const hipDoubleComplex *csrValA, const int *csrColIndA, const int *csrRowPtrA, int nnzA, const int *nnzPerRow, hipDoubleComplex *csrValC, int *csrColIndC, int *csrRowPtrC, hipDoubleComplex tol)

Convert a sparse CSR matrix into a compressed sparse CSR matrix.

hipsparseXcsr2csr_compress converts a CSR matrix into a compressed CSR matrix by removing entries in the input CSR matrix that are below a non-negative threshold tol:

\[ C(i,j) = A(i, j) \text{ if |A(i, j)| > tol} \]

The user must first call nnz_compress to determine the number of nonzeros per row as well as the total number of nonzeros that will exist in resulting compressed CSR matrix. The user then uses this information to allocate the column indices array csrColIndC and the values array csrValC. The user then calls hipsparseXcsr2csr_compress to complete the conversion.

Example

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

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

// Sparse matrix in CSR format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtrA[4] = {0, 3, 5, 8};
int hcsrColIndA[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcsrValA[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m    = 3;
int n    = 5;
int nnzA = 8;

float tol = 5.9f;

int* dcsrRowPtrA = nullptr;
int* dcsrColIndA = nullptr;
float* dcsrValA = nullptr;
hipMalloc((void**)&dcsrRowPtrA, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColIndA, sizeof(int) * nnzA);
hipMalloc((void**)&dcsrValA, sizeof(float) * nnzA);

hipMemcpy(dcsrRowPtrA, hcsrRowPtrA, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColIndA, hcsrColIndA, sizeof(int) * nnzA, hipMemcpyHostToDevice);
hipMemcpy(dcsrValA, hcsrValA, sizeof(float) * nnzA, hipMemcpyHostToDevice);

// Allocate memory for the nnz_per_row array
int* dnnz_per_row;
hipMalloc((void**)&dnnz_per_row, sizeof(int) * m);

// Call snnz_compress() which fills in nnz_per_row array and finds the number
// of entries that will be in the compressed CSR matrix
int nnzC;
hipsparseSnnz_compress(handle, m, descr, dcsrValA, dcsrRowPtrA, dnnz_per_row, &nnzC, tol);

int* dcsrRowPtrC = nullptr;
int* dcsrColIndC = nullptr;
float* dcsrValC = nullptr;
hipMalloc((void**)&dcsrRowPtrC, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColIndC, sizeof(int) * nnzC);
hipMalloc((void**)&dcsrValC, sizeof(float) * nnzC);

hipsparseScsr2csr_compress(handle,
                           m,
                           n,
                           descr,
                           dcsrValA,
                           dcsrColIndA,
                           dcsrRowPtrA,
                           nnzA,
                           dnnz_per_row,
                           dcsrValC,
                           dcsrColIndC,
                           dcsrRowPtrC,
                           tol);

hipFree(dcsrRowPtrA);
hipFree(dcsrColIndA);
hipFree(dcsrValA);

hipFree(dcsrRowPtrC);
hipFree(dcsrColIndC);
hipFree(dcsrValC);

hipFree(dnnz_per_row);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

In the case of complex matrices only the magnitude of the real part of tol is used.

hipsparseXpruneCsr2csr_bufferSize()

hipsparseStatus_t hipsparseSpruneCsr2csr_bufferSize(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const float *threshold, const hipsparseMatDescr_t descrC, const float *csrValC, const int *csrRowPtrC, const int *csrColIndC, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneCsr2csr_bufferSize(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const double *threshold, const hipsparseMatDescr_t descrC, const double *csrValC, const int *csrRowPtrC, const int *csrColIndC, size_t *pBufferSizeInBytes)

Convert and prune sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csr_bufferSize returns the size of the temporary buffer that is required by hipsparseXpruneCsr2csrNnz and hipsparseXpruneCsr2csr. The temporary storage buffer must be allocated by the user.

hipsparseXpruneCsr2csr_bufferSizeExt()

hipsparseStatus_t hipsparseSpruneCsr2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const float *threshold, const hipsparseMatDescr_t descrC, const float *csrValC, const int *csrRowPtrC, const int *csrColIndC, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneCsr2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const double *threshold, const hipsparseMatDescr_t descrC, const double *csrValC, const int *csrRowPtrC, const int *csrColIndC, size_t *pBufferSizeInBytes)

Convert and prune sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csr_bufferSizeExt returns the size of the temporary buffer that is required by hipsparseXpruneCsr2csrNnz and hipsparseXpruneCsr2csr. The temporary storage buffer must be allocated by the user.

hipsparseXpruneCsr2csrNnz()

hipsparseStatus_t hipsparseSpruneCsr2csrNnz(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const float *threshold, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr, void *buffer)

hipsparseStatus_t hipsparseDpruneCsr2csrNnz(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const double *threshold, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr, void *buffer)

Convert and prune sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csrNnz computes the number of nonzero elements per row and the total number of nonzero elements in a sparse CSR matrix once elements less than the threshold are pruned from the matrix.

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXpruneCsr2csr()

hipsparseStatus_t hipsparseSpruneCsr2csr(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, const float *threshold, const hipsparseMatDescr_t descrC, float *csrValC, const int *csrRowPtrC, int *csrColIndC, void *buffer)

hipsparseStatus_t hipsparseDpruneCsr2csr(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, const double *threshold, const hipsparseMatDescr_t descrC, double *csrValC, const int *csrRowPtrC, int *csrColIndC, void *buffer)

Convert and prune sparse CSR matrix into a sparse CSR matrix.

This function converts the sparse CSR matrix A into a sparse CSR matrix C by pruning values in A that are less than the threshold. All the parameters are assumed to have been pre-allocated by the user. The user first calls hipsparseXpruneCsr2csr_bufferSize() to determine the size of the buffer used by hipsparseXpruneCsr2csrNnz() and hipsparseXpruneCsr2csr() which the user then allocates. The user then allocates csrRowPtrC to have m+1 elements and then calls hipsparseXpruneCsr2csrNnz() which fills in the csrRowPtrC array stores then number of elements that are larger than the pruning threshold in nnzTotalDevHostPtr. The user then calls hipsparseXpruneCsr2csr() to complete the conversion. It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXpruneCsr2csrByPercentage_bufferSize()

hipsparseStatus_t hipsparseSpruneCsr2csrByPercentage_bufferSize(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, float percentage, const hipsparseMatDescr_t descrC, const float *csrValC, const int *csrRowPtrC, const int *csrColIndC, pruneInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneCsr2csrByPercentage_bufferSize(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, double percentage, const hipsparseMatDescr_t descrC, const double *csrValC, const int *csrRowPtrC, const int *csrColIndC, pruneInfo_t info, size_t *pBufferSizeInBytes)

Convert and prune by percentage a sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csrByPercentage_bufferSize returns the size of the temporary buffer that is required by hipsparseXpruneCsr2csrNnzByPercentage. The temporary storage buffer must be allocated by the user.

hipsparseXpruneCsr2csrByPercentage_bufferSizeExt()

hipsparseStatus_t hipsparseSpruneCsr2csrByPercentage_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, float percentage, const hipsparseMatDescr_t descrC, const float *csrValC, const int *csrRowPtrC, const int *csrColIndC, pruneInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDpruneCsr2csrByPercentage_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, double percentage, const hipsparseMatDescr_t descrC, const double *csrValC, const int *csrRowPtrC, const int *csrColIndC, pruneInfo_t info, size_t *pBufferSizeInBytes)

Convert and prune by percentage a sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csrByPercentage_bufferSizeExt returns the size of the temporary buffer that is required by hipsparseXpruneCsr2csrNnzByPercentage. The temporary storage buffer must be allocated by the user.

hipsparseXpruneCsr2csrNnzByPercentage()

hipsparseStatus_t hipsparseSpruneCsr2csrNnzByPercentage(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, float percentage, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr, pruneInfo_t info, void *buffer)

hipsparseStatus_t hipsparseDpruneCsr2csrNnzByPercentage(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, double percentage, const hipsparseMatDescr_t descrC, int *csrRowPtrC, int *nnzTotalDevHostPtr, pruneInfo_t info, void *buffer)

Convert and prune by percentage a sparse CSR matrix into a sparse CSR matrix.

hipsparseXpruneCsr2csrNnzByPercentage computes the number of nonzero elements per row and the total number of nonzero elements in a sparse CSR matrix once elements less than the threshold are pruned from the matrix.

Note

The routine does support asynchronous execution if the pointer mode is set to device.

hipsparseXpruneCsr2csrByPercentage()

hipsparseStatus_t hipsparseSpruneCsr2csrByPercentage(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const float *csrValA, const int *csrRowPtrA, const int *csrColIndA, float percentage, const hipsparseMatDescr_t descrC, float *csrValC, const int *csrRowPtrC, int *csrColIndC, pruneInfo_t info, void *buffer)

hipsparseStatus_t hipsparseDpruneCsr2csrByPercentage(hipsparseHandle_t handle, int m, int n, int nnzA, const hipsparseMatDescr_t descrA, const double *csrValA, const int *csrRowPtrA, const int *csrColIndA, double percentage, const hipsparseMatDescr_t descrC, double *csrValC, const int *csrRowPtrC, int *csrColIndC, pruneInfo_t info, void *buffer)

Convert and prune by percentage a sparse CSR matrix into a sparse CSR matrix.

This function converts the sparse CSR matrix A into a sparse CSR matrix C by pruning values in A that are less than the threshold. All the parameters are assumed to have been pre-allocated by the user. The user first calls hipsparseXpruneCsr2csr_bufferSize() to determine the size of the buffer used by hipsparseXpruneCsr2csrNnz() and hipsparseXpruneCsr2csr() which the user then allocates. The user then allocates csrRowPtrC to have m+1 elements and then calls hipsparseXpruneCsr2csrNnz() which fills in the csrRowPtrC array stores then number of elements that are larger than the pruning threshold in nnzTotalDevHostPtr. The user then calls hipsparseXpruneCsr2csr() to complete the conversion. It is executed asynchronously with respect to the host and may return control to the application on the host before the entire result is ready.

hipsparseXhyb2csr()

hipsparseStatus_t hipsparseShyb2csr(hipsparseHandle_t handle, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, float *csrSortedValA, int *csrSortedRowPtrA, int *csrSortedColIndA)

hipsparseStatus_t hipsparseDhyb2csr(hipsparseHandle_t handle, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, double *csrSortedValA, int *csrSortedRowPtrA, int *csrSortedColIndA)

hipsparseStatus_t hipsparseChyb2csr(hipsparseHandle_t handle, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, hipComplex *csrSortedValA, int *csrSortedRowPtrA, int *csrSortedColIndA)

hipsparseStatus_t hipsparseZhyb2csr(hipsparseHandle_t handle, const hipsparseMatDescr_t descrA, const hipsparseHybMat_t hybA, hipDoubleComplex *csrSortedValA, int *csrSortedRowPtrA, int *csrSortedColIndA)

Convert a sparse HYB matrix into a sparse CSR matrix.

hipsparseXhyb2csr converts a HYB matrix into a CSR matrix.

Note

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

hipsparseXcoo2csr()

hipsparseStatus_t hipsparseXcoo2csr(hipsparseHandle_t handle, const int *cooRowInd, int nnz, int m, int *csrRowPtr, hipsparseIndexBase_t idxBase)

Convert a sparse COO matrix into a sparse CSR matrix.

hipsparseXcoo2csr converts the COO array containing the row indices into a CSR array of row offsets, that point to the start of every row. It is assumed that the COO row index array is sorted and that all arrays have been allocated prior to calling hipsparseXcoo2csr.

For example, given the COO row indices array:

\[ \begin{align} \text{cooRowInd} &= \begin{bmatrix} 0 & 0 & 1 & 2 & 2 & 4 & 4 & 4 \end{bmatrix} \end{align} \]

the resulting CSR row pointer array after calling hipsparseXcoo2csr is:

\[ \begin{align} \text{csrRowPtr} &= \begin{bmatrix} 0 & 2 & 3 & 5 & 8 \end{bmatrix} \end{align} \]

Example

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

// Sparse matrix in COO format
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcooRowInd[8] = {0, 0, 0, 1, 1, 2, 2, 2};
int hcooColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcooVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;

int* dcooRowInd = nullptr;
int* dcooColInd = nullptr;
hipMalloc((void**)&dcooRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcooColInd, sizeof(int) * nnz);

hipMemcpy(dcooRowInd, hcooRowInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcooColInd, hcooColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);

int* dcsrRowPtr = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));

hipsparseXcoo2csr(handle, dcooRowInd, nnz, m, dcsrRowPtr, base);

hipFree(dcooRowInd);
hipFree(dcooColInd);

hipFree(dcsrRowPtr);

hipsparseDestroy(handle);

Note

It can also be used, to convert a COO array containing the column indices into a CSC array of column offsets, that point to the start of every column. Then, it is assumed that the COO column index array is sorted, instead.

Note

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

hipsparseCreateIdentityPermutation()

hipsparseStatus_t hipsparseCreateIdentityPermutation(hipsparseHandle_t handle, int n, int *p)

Create the identity map.

hipsparseCreateIdentityPermutation stores the identity map in p, such that \(p = 0:1:(n-1)\).

for(i = 0; i < n; ++i)
{
    p[i] = i;
}

Example

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

int n = 10;

int* dperm = nullptr;
hipMalloc((void**)&dperm, sizeof(int) * n);

hipsparseCreateIdentityPermutation(handle, n, dperm);

hipsparseDestroy(handle);

Note

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

hipsparseXcsrsort_bufferSizeExt()

hipsparseStatus_t hipsparseXcsrsort_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, const int *csrRowPtr, const int *csrColInd, size_t *pBufferSizeInBytes)

Sort a sparse CSR matrix.

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

hipsparseXcsrsort()

hipsparseStatus_t hipsparseXcsrsort(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, const int *csrRowPtr, int *csrColInd, int *P, void *pBuffer)

Sort a sparse CSR matrix.

hipsparseXcsrsort sorts a matrix in CSR format. The sorted permutation vector perm can be used to obtain sorted csrVal array. In this case, perm must be initialized as the identity permutation, see hipsparseCreateIdentityPermutation(). To apply the permutation vector to the CSR values, see hipsparse hipsparseSgthr().

hipsparseXcsrsort requires extra temporary storage buffer that has to be allocated by the user. Storage buffer size can be determined by hipsparseXcsrsort_bufferSizeExt().

Example

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

hipsparseMatDescr_t descr;
hipsparseCreateMatDescr(&descr);

// Sparse matrix in CSR format (columns unsorted)
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcsrRowPtr[4] = {0, 3, 5, 8};
int hcsrColInd[8] = {3, 1, 0, 2, 1, 0, 4, 3};
float hcsrVal[8]  = {3.0f, 2.0f, 1.0f, 5.0f, 4.0f, 6.0f, 8.0f, 7.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;

int* dcsrRowPtr = nullptr;
int* dcsrColInd = nullptr;
float* dcsrVal = nullptr;
hipMalloc((void**)&dcsrRowPtr, sizeof(int) * (m + 1));
hipMalloc((void**)&dcsrColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcsrVal, sizeof(float) * nnz);

hipMemcpy(dcsrRowPtr, hcsrRowPtr, sizeof(int) * (m + 1), hipMemcpyHostToDevice);
hipMemcpy(dcsrColInd, hcsrColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcsrVal, hcsrVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

size_t bufferSize;
hipsparseXcsrsort_bufferSizeExt(handle,
                                m,
                                n,
                                nnz,
                                dcsrRowPtr,
                                dcsrColInd,
                                &bufferSize);

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

int* dperm = nullptr;
hipMalloc((void**)&dperm, sizeof(int) * nnz);
hipsparseCreateIdentityPermutation(handle, nnz, dperm);

hipsparseXcsrsort(handle,
                  m,
                  n,
                  nnz,
                  descr,
                  dcsrRowPtr,
                  dcsrColInd,
                  dperm,
                  dbuffer);

float* dcsrValSorted = nullptr;
hipMalloc((void**)&dcsrValSorted, sizeof(float) * nnz);
hipsparseSgthr(handle, nnz, dcsrVal, dcsrValSorted, dperm, HIPSPARSE_INDEX_BASE_ZERO);

hipFree(dcsrRowPtr);
hipFree(dcsrColInd);
hipFree(dcsrVal);
hipFree(dcsrValSorted);

hipFree(dbuffer);
hipFree(dperm);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

perm can be NULL if a sorted permutation vector is not required.

Note

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

hipsparseXcscsort_bufferSizeExt()

hipsparseStatus_t hipsparseXcscsort_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, const int *cscColPtr, const int *cscRowInd, size_t *pBufferSizeInBytes)

Sort a sparse CSC matrix.

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

hipsparseXcscsort()

hipsparseStatus_t hipsparseXcscsort(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, const int *cscColPtr, int *cscRowInd, int *P, void *pBuffer)

Sort a sparse CSC matrix.

hipsparseXcscsort sorts a matrix in CSC format. The sorted permutation vector perm can be used to obtain sorted csc_val array. In this case, perm must be initialized as the identity permutation, see hipsparseCreateIdentityPermutation(). To apply the permutation vector to the CSC values, see hipsparse hipsparseSgthr().

hipsparseXcscsort requires extra temporary storage buffer that has to be allocated by the user. Storage buffer size can be determined by hipsparseXcscsort_bufferSizeExt().

Example

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

hipsparseMatDescr_t descr;
hipsparseCreateMatDescr(&descr);

// Sparse matrix in CSC format (unsorted row indices)
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcscRowInd[8] = {2, 0, 1, 0, 1, 2, 0, 2};
int hcscColPtr[6] = {0, 2, 4, 5, 7, 8};
float hcscVal[8]  = {6.0f, 1.0f, 4.0f, 2.0f, 5.0f, 7.0f, 3.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;

int* dcscRowInd = nullptr;
int* dcscColPtr = nullptr;
float* dcscVal = nullptr;
hipMalloc((void**)&dcscRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcscColPtr, sizeof(int) * (n + 1));
hipMalloc((void**)&dcscVal, sizeof(float) * nnz);

hipMemcpy(dcscRowInd, hcscRowInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcscColPtr, hcscColPtr, sizeof(int) * (n + 1), hipMemcpyHostToDevice);
hipMemcpy(dcscVal, hcscVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

size_t bufferSize;
hipsparseXcscsort_bufferSizeExt(handle,
                                m,
                                n,
                                nnz,
                                dcscColPtr,
                                dcscRowInd,
                                &bufferSize);

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

int* dperm = nullptr;
hipMalloc((void**)&dperm, sizeof(int) * nnz);
hipsparseCreateIdentityPermutation(handle, nnz, dperm);

hipsparseXcscsort(handle,
                  m,
                  n,
                  nnz,
                  descr,
                  dcscColPtr,
                  dcscRowInd,
                  dperm,
                  dbuffer);

float* dcscValSorted = nullptr;
hipMalloc((void**)&dcscValSorted, sizeof(float) * nnz);
hipsparseSgthr(handle, nnz, dcscVal, dcscValSorted, dperm, HIPSPARSE_INDEX_BASE_ZERO);

hipFree(dcscRowInd);
hipFree(dcscColPtr);
hipFree(dcscVal);
hipFree(dcscValSorted);

hipFree(dbuffer);
hipFree(dperm);

hipsparseDestroyMatDescr(descr);
hipsparseDestroy(handle);

Note

perm can be NULL if a sorted permutation vector is not required.

Note

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

hipsparseXcoosort_bufferSizeExt()

hipsparseStatus_t hipsparseXcoosort_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, const int *cooRows, const int *cooCols, size_t *pBufferSizeInBytes)

Sort a sparse COO matrix.

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

hipsparseXcoosortByRow()

hipsparseStatus_t hipsparseXcoosortByRow(hipsparseHandle_t handle, int m, int n, int nnz, int *cooRows, int *cooCols, int *P, void *pBuffer)

Sort a sparse COO matrix by row.

hipsparseXcoosortByRow sorts a matrix in COO format by row. The sorted permutation vector perm can be used to obtain sorted cooVal array. In this case, perm must be initialized as the identity permutation, see hipsparseCreateIdentityPermutation(). To apply the permutation vector to the COO values, see hipsparse hipsparseSgthr().

hipsparseXcoosortByRow requires extra temporary storage buffer that has to be allocated by the user. Storage buffer size can be determined by hipsparseXcoosort_bufferSizeExt().

Example

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

// Sparse matrix in COO format (with unsorted row indices)
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcooRowInd[8] = {0, 2, 0, 1, 1, 0, 2, 2};
int hcooColInd[8] = {0, 0, 1, 1, 2, 3, 3, 4};
float hcooVal[8]   = {1.0f, 6.0f, 2.0f, 4.0f, 5.0f, 3.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;

int* dcooRowInd = nullptr;
int* dcooColInd = nullptr;
float* dcooVal = nullptr;
hipMalloc((void**)&dcooRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcooColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcooVal, sizeof(float) * nnz);

hipMemcpy(dcooRowInd, hcooRowInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcooColInd, hcooColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcooVal, hcooVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

size_t bufferSize;
hipsparseXcoosort_bufferSizeExt(handle, m, n, nnz, dcooRowInd, dcooColInd, &bufferSize);

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

int* dperm = nullptr;
hipMalloc((void**)&dperm, sizeof(int) * nnz);
hipsparseCreateIdentityPermutation(handle, nnz, dperm);

hipsparseXcoosortByRow(handle, m, n, nnz, dcooRowInd, dcooColInd, dperm, dbuffer);

float* dcooValSorted = nullptr;
hipMalloc((void**)&dcooValSorted, sizeof(float) * nnz);
hipsparseSgthr(handle, nnz, dcooVal, dcooValSorted, dperm, base);

hipFree(dcooRowInd);
hipFree(dcooColInd);
hipFree(dcooVal);
hipFree(dcooValSorted);
hipFree(dperm);

hipFree(dbuffer);

hipsparseDestroy(handle);

Note

perm can be NULL if a sorted permutation vector is not required.

Note

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

hipsparseXcoosortByColumn()

hipsparseStatus_t hipsparseXcoosortByColumn(hipsparseHandle_t handle, int m, int n, int nnz, int *cooRows, int *cooCols, int *P, void *pBuffer)

Sort a sparse COO matrix by column.

hipsparseXcoosortByColumn sorts a matrix in COO format by column. The sorted permutation vector perm can be used to obtain sorted cooVal array. In this case, perm must be initialized as the identity permutation, see hipsparseCreateIdentityPermutation(). To apply the permutation vector to the COO values, see hipsparse hipsparseSgthr().

hipsparseXcoosortByColumn requires extra temporary storage buffer that has to be allocated by the user. Storage buffer size can be determined by hipsparseXcoosort_bufferSizeExt().

Example

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

// Sparse matrix in COO format (with unsorted column indices)
//     1 2 0 3 0
// A = 0 4 5 0 0
//     6 0 0 7 8
int hcooRowInd[8] = {0, 0, 0, 1, 1, 2, 2, 2};
int hcooColInd[8] = {0, 1, 3, 1, 2, 0, 3, 4};
float hcooVal[8]   = {1.0f, 2.0f, 3.0f, 4.0f, 5.0f, 6.0f, 7.0f, 8.0f}; 

int m         = 3;
int n         = 5;
int nnz       = 8;
hipsparseIndexBase_t base = HIPSPARSE_INDEX_BASE_ZERO;

int* dcooRowInd = nullptr;
int* dcooColInd = nullptr;
float* dcooVal = nullptr;
hipMalloc((void**)&dcooRowInd, sizeof(int) * nnz);
hipMalloc((void**)&dcooColInd, sizeof(int) * nnz);
hipMalloc((void**)&dcooVal, sizeof(float) * nnz);

hipMemcpy(dcooRowInd, hcooRowInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcooColInd, hcooColInd, sizeof(int) * nnz, hipMemcpyHostToDevice);
hipMemcpy(dcooVal, hcooVal, sizeof(float) * nnz, hipMemcpyHostToDevice);

size_t bufferSize;
hipsparseXcoosort_bufferSizeExt(handle, m, n, nnz, dcooRowInd, dcooColInd, &bufferSize);

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

int* dperm = nullptr;
hipMalloc((void**)&dperm, sizeof(int) * nnz);
hipsparseCreateIdentityPermutation(handle, nnz, dperm);

hipsparseXcoosortByColumn(handle, m, n, nnz, dcooRowInd, dcooColInd, dperm, dbuffer);

float* dcooValSorted = nullptr;
hipMalloc((void**)&dcooValSorted, sizeof(float) * nnz);
hipsparseSgthr(handle, nnz, dcooVal, dcooValSorted, dperm, base);

hipFree(dcooRowInd);
hipFree(dcooColInd);
hipFree(dcooVal);
hipFree(dcooValSorted);
hipFree(dperm);

hipFree(dbuffer);

hipsparseDestroy(handle);

Note

perm can be NULL if a sorted permutation vector is not required.

Note

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

hipsparseXgebsr2gebsr_bufferSize()

hipsparseStatus_t hipsparseSgebsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const float *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, int rowBlockDimC, int colBlockDimC, int *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDgebsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const double *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, int rowBlockDimC, int colBlockDimC, int *pBufferSizeInBytes)

hipsparseStatus_t hipsparseCgebsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const hipComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, int rowBlockDimC, int colBlockDimC, int *pBufferSizeInBytes)

hipsparseStatus_t hipsparseZgebsr2gebsr_bufferSize(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, int rowBlockDimC, int colBlockDimC, int *pBufferSizeInBytes)

This function computes the the size of the user allocated temporary storage buffer used when converting a sparse GEBSR matrix to another sparse GEBSR matrix.

hipsparseXgebsr2gebsr_bufferSize returns the size of the temporary storage buffer that is required by hipsparseXgebsr2gebsrNnz() and hipsparseXgebsr2gebsr(). The temporary storage buffer must be allocated by the user.

hipsparseXgebsr2gebsrNnz()

hipsparseStatus_t hipsparseXgebsr2gebsrNnz(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, const hipsparseMatDescr_t descrC, int *bsrRowPtrC, int rowBlockDimC, int colBlockDimC, int *nnzTotalDevHostPtr, void *buffer)

This function is used when converting a GEBSR sparse matrix A to another GEBSR sparse matrix C. Specifically, this function determines the number of non-zero blocks that will exist in C (stored using either a host or device pointer), and computes the row pointer array for C.

The routine does support asynchronous execution.

hipsparseXgebsr2gebsr()

hipsparseStatus_t hipsparseSgebsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const float *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, const hipsparseMatDescr_t descrC, float *bsrValC, int *bsrRowPtrC, int *bsrColIndC, int rowBlockDimC, int colBlockDimC, void *buffer)

hipsparseStatus_t hipsparseDgebsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const double *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, const hipsparseMatDescr_t descrC, double *bsrValC, int *bsrRowPtrC, int *bsrColIndC, int rowBlockDimC, int colBlockDimC, void *buffer)

hipsparseStatus_t hipsparseCgebsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const hipComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, const hipsparseMatDescr_t descrC, hipComplex *bsrValC, int *bsrRowPtrC, int *bsrColIndC, int rowBlockDimC, int colBlockDimC, void *buffer)

hipsparseStatus_t hipsparseZgebsr2gebsr(hipsparseHandle_t handle, hipsparseDirection_t dirA, int mb, int nb, int nnzb, const hipsparseMatDescr_t descrA, const hipDoubleComplex *bsrValA, const int *bsrRowPtrA, const int *bsrColIndA, int rowBlockDimA, int colBlockDimA, const hipsparseMatDescr_t descrC, hipDoubleComplex *bsrValC, int *bsrRowPtrC, int *bsrColIndC, int rowBlockDimC, int colBlockDimC, void *buffer)

This function converts the GEBSR sparse matrix A to another GEBSR sparse matrix C.

The conversion uses three steps. First, the user calls hipsparseXgebsr2gebsr_bufferSize() to determine the size of the required temporary storage buffer. The user then allocates this buffer. Secondly, the user then allocates mbC+1 integers for the row pointer array for C where:

\[\begin{split} \begin{align} \text{mbC} &= \text{(m - 1) / rowBlockDimC + 1} \\ \text{nbC} &= \text{(n - 1) / colBlockDimC + 1} \end{align} \end{split}\]

The user then calls hipsparseXgebsr2gebsrNnz() to fill in the row pointer array for C ( bsrRowPtrC ) and determine the number of non-zero blocks that will exist in C. Finally, the user allocates space for the column indices array of C to have nnzbC elements and space for the values array of C to have nnzbC*rowBlockDimC*colBlockDimC and then calls hipsparseXgebsr2gebsr() to complete the conversion.

It may be the case that rowBlockDimC does not divide evenly into m and/or colBlockDim does not divide evenly into n. In these cases, the GEBSR matrix is expanded in size in order to fit full GEBSR blocks. For example, if the original GEBSR matrix A (using rowBlockDimA=2, colBlockDimA=3) looks like:

\[\begin{split} \left[ \begin{array}{c | c} \begin{array}{c c c} 1 & 0 & 0 \\ 3 & 4 & 0 \end{array} & \begin{array}{c c c} 2 & 0 & 0 \\ 4 & 5 & 6 \end{array} \\ \hline \begin{array}{c c c} 1 & 2 & 3 \\ 1 & 2 & 0 \end{array} & \begin{array}{c c c} 4 & 0 & 0 \\ 3 & 0 & 1 \end{array} \\ \end{array} \right] \end{split}\]

then if we specify rowBlockDimC=3 and colBlockDimC=2, our output GEBSR matrix C would be:

\[\begin{split} \left[ \begin{array}{c | c | c} \begin{array}{c c} 1 & 0 \\ 3 & 4 \\ 1 & 2 \end{array} & \begin{array}{c c} 0 & 2 \\ 0 & 4 \\ 3 & 4 \end{array} & \begin{array}{c c} 0 & 0 \\ 5 & 6 \\ 0 & 0 \end{array} \\ \hline \begin{array}{c c} 1 & 2 \\ 0 & 0 \\ 0 & 0 \end{array} & \begin{array}{c c} 0 & 3 \\ 0 & 0 \\ 0 & 0 \end{array} & \begin{array}{c c} 0 & 1 \\ 0 & 0 \\ 0 & 0 \end{array} \\ \end{array} \right] \end{split}\]

Example

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

hipsparseMatDescr_t descrA;
hipsparseCreateMatDescr(&descrA);

hipsparseMatDescr_t descrC;
hipsparseCreateMatDescr(&descrC);

// Sparse matrix in BSR format
//     1 2 | 0 3 | 0 0 
//     0 4 | 5 0 | 0 1
// A = 6 0 | 0 7 | 8 0
//     --------------- 
//     0 0 | 3 0 | 2 2
//     1 0 | 0 0 | 4 3 
//     7 2 | 0 0 | 1 4
int hbsrRowPtrA[3] = {0, 3, 6};
int hbsrColIndA[6] = {0, 1, 2, 0, 1, 2};
float hbsrValA[36]  = {1.0f, 2.0f, 0.0f, 4.0f, 6.0f, 0.0f, 
                       0.0f, 3.0f, 5.0f, 0.0f, 0.0f, 7.0f, 
                       0.0f, 0.0f, 0.0f, 1.0f, 8.0f, 0.0f, 
                       0.0f, 0.0f, 1.0f, 0.0f, 7.0f, 2.0f,
                       3.0f, 0.0f, 0.0f, 0.0f, 0.0f, 0.0f,
                       2.0f, 2.0f, 4.0f, 3.0f, 1.0f, 4.0f}; 

int m           = 6;
int n           = 6;
int rowBlockDimA = 3;
int colBlockDimA = 2;
int rowBlockDimC = 2;
int colBlockDimC = 2;
hipsparseDirection_t dirA = HIPSPARSE_DIRECTION_ROW;

int mbA   = (m + rowBlockDimA - 1) / rowBlockDimA;
int nbA   = (n + colBlockDimA - 1) / colBlockDimA;
int nnzbA = 6;

int mbC   = (m + rowBlockDimC - 1) / rowBlockDimC;
int nbC   = (n + colBlockDimC - 1) / colBlockDimC;

int* dbsrRowPtrA = nullptr;
int* dbsrColIndA = nullptr;
float* dbsrValA = nullptr;
hipMalloc((void**)&dbsrRowPtrA, sizeof(int) * (mbA + 1));
hipMalloc((void**)&dbsrColIndA, sizeof(int) * nnzbA);
hipMalloc((void**)&dbsrValA, sizeof(float) * rowBlockDimA * colBlockDimA * nnzbA);

hipMemcpy(dbsrRowPtrA, hbsrRowPtrA, sizeof(int) * (mbA + 1), hipMemcpyHostToDevice);
hipMemcpy(dbsrColIndA, hbsrColIndA, sizeof(int) * nnzbA, hipMemcpyHostToDevice);
hipMemcpy(dbsrValA, hbsrValA, sizeof(float) * rowBlockDimA * colBlockDimA * nnzbA, hipMemcpyHostToDevice);

int* dbsrRowPtrC = nullptr;
hipMalloc((void**)&dbsrRowPtrC, sizeof(int) * (mbC + 1));

size_t bufferSize;
hipsparseSgebsr2gebsr_bufferSize(handle, 
                                 dirA, 
                                 mbA, 
                                 nbA, 
                                 nnzbA, 
                                 descrA, 
                                 dbsrValA, 
                                 dbsrRowPtrA, 
                                 dbsrColIndA, 
                                 rowBlockDimA, 
                                 colBlockDimA, 
                                 rowBlockDimC, 
                                 colBlockDimC, 
                                 &bufferSize);

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

int nnzbC;
hipsparseXgebsr2gebsrNnz(handle,
                         dirA,
                         mbA,
                         nbA,
                         nnzbA,
                         descrA,
                         dbsrRowPtrA,
                         dbsrColIndA,
                         rowBlockDimA,
                         colBlockDimA,
                         descrC,
                         dbsrRowPtrC,
                         rowBlockDimC,
                         colBlockDimC,
                         &nnzbC,
                         dbuffer);

hipDeviceSynchronize();

int* dbsrColIndC = nullptr;
float* dbsrValC = nullptr;
hipMalloc((void**)&dbsrColIndC, sizeof(int) * nnzbC);
hipMalloc((void**)&dbsrValC, sizeof(float) * rowBlockDimC * colBlockDimC * nnzbC);

hipsparseSgebsr2gebsr(handle, 
                      dirA, 
                      mbA, 
                      nbA, 
                      nnzbA, 
                      descrA, 
                      dbsrValA, 
                      dbsrRowPtrA, 
                      dbsrColIndA, 
                      rowBlockDimA, 
                      colBlockDimA, 
                      descrC, 
                      dbsrValC, 
                      dbsrRowPtrC, 
                      dbsrColIndC, 
                      rowBlockDimC, 
                      colBlockDimC, 
                      dbuffer);

hipFree(dbsrRowPtrA);
hipFree(dbsrColIndA);
hipFree(dbsrValA);

hipFree(dbsrRowPtrC);
hipFree(dbsrColIndC);
hipFree(dbsrValC);

hipFree(dbuffer);

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

hipsparseXcsru2csr_bufferSizeExt()

hipsparseStatus_t hipsparseScsru2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, float *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseDcsru2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, double *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseCcsru2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, hipComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, size_t *pBufferSizeInBytes)

hipsparseStatus_t hipsparseZcsru2csr_bufferSizeExt(hipsparseHandle_t handle, int m, int n, int nnz, hipDoubleComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, size_t *pBufferSizeInBytes)

This function calculates the amount of temporary storage in bytes required for hipsparseXcsru2csr() and hipsparseXcsr2csru().

hipsparseXcsru2csr()

hipsparseStatus_t hipsparseScsru2csr(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, float *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseDcsru2csr(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, double *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseCcsru2csr(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, hipComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseZcsru2csr(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, hipDoubleComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

This function converts unsorted CSR format to sorted CSR format. The required temporary storage has to be allocated by the user.

hipsparseXcsr2csru()

hipsparseStatus_t hipsparseScsr2csru(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, float *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseDcsr2csru(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, double *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseCcsr2csru(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, hipComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

hipsparseStatus_t hipsparseZcsr2csru(hipsparseHandle_t handle, int m, int n, int nnz, const hipsparseMatDescr_t descrA, hipDoubleComplex *csrVal, const int *csrRowPtr, int *csrColInd, csru2csrInfo_t info, void *pBuffer)

This function converts sorted CSR format to unsorted CSR format. The required temporary storage has to be allocated by the user.