Skip to content

Commit

Permalink
Merge pull request ROCm#1282 from emankov/HIPIFY
Browse files Browse the repository at this point in the history
[HIPIFY][ROCm#674][rocSPARSE][feature] rocSPARSE support - Step 98 - function `rocsparse_sparse_to_dense`
  • Loading branch information
emankov authored Jan 16, 2024
2 parents d8188fa + 4c282f4 commit 5d82b98
Show file tree
Hide file tree
Showing 9 changed files with 150 additions and 7 deletions.
1 change: 1 addition & 0 deletions bin/hipify-perl
Original file line number Diff line number Diff line change
Expand Up @@ -2610,6 +2610,7 @@ sub rocSubstitutions {
subst("cusparseSpVecGetIndexBase", "rocsparse_spvec_get_index_base", "library");
subst("cusparseSpVecGetValues", "rocsparse_spvec_get_values", "library");
subst("cusparseSpVecSetValues", "rocsparse_spvec_set_values", "library");
subst("cusparseSparseToDense", "rocsparse_sparse_to_dense", "library");
subst("cusparseSpruneCsr2csr", "rocsparse_sprune_csr2csr", "library");
subst("cusparseSpruneCsr2csrByPercentage", "rocsparse_sprune_csr2csr_by_percentage", "library");
subst("cusparseSpruneCsr2csrByPercentage_bufferSizeExt", "rocsparse_sprune_csr2csr_by_percentage_buffer_size", "library");
Expand Down
2 changes: 1 addition & 1 deletion docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md
Original file line number Diff line number Diff line change
Expand Up @@ -910,7 +910,7 @@
|`cusparseSpVecGetIndexBase`|10.2| |12.0| |`hipsparseSpVecGetIndexBase`|4.1.0| |6.0.0| | |`rocsparse_spvec_get_index_base`|4.1.0| |6.0.0| | |
|`cusparseSpVecGetValues`|10.2| | | |`hipsparseSpVecGetValues`|4.1.0| | | | |`rocsparse_spvec_get_values`|4.1.0| | | | |
|`cusparseSpVecSetValues`|10.2| | | |`hipsparseSpVecSetValues`|4.1.0| | | | |`rocsparse_spvec_set_values`|4.1.0| | | | |
|`cusparseSparseToDense`|11.1| |12.0| |`hipsparseSparseToDense`|4.2.0| |6.0.0| | | | | | | | |
|`cusparseSparseToDense`|11.1| |12.0| |`hipsparseSparseToDense`|4.2.0| |6.0.0| | |`rocsparse_sparse_to_dense`|4.1.0| |6.0.0| | |
|`cusparseSparseToDense_bufferSize`|11.1| |12.0| |`hipsparseSparseToDense_bufferSize`|4.2.0| |6.0.0| | | | | | | | |


Expand Down
2 changes: 1 addition & 1 deletion docs/tables/CUSPARSE_API_supported_by_ROC.md
Original file line number Diff line number Diff line change
Expand Up @@ -910,7 +910,7 @@
|`cusparseSpVecGetIndexBase`|10.2| |12.0| |`rocsparse_spvec_get_index_base`|4.1.0| |6.0.0| | |
|`cusparseSpVecGetValues`|10.2| | | |`rocsparse_spvec_get_values`|4.1.0| | | | |
|`cusparseSpVecSetValues`|10.2| | | |`rocsparse_spvec_set_values`|4.1.0| | | | |
|`cusparseSparseToDense`|11.1| |12.0| | | | | | | |
|`cusparseSparseToDense`|11.1| |12.0| |`rocsparse_sparse_to_dense`|4.1.0| |6.0.0| | |
|`cusparseSparseToDense_bufferSize`|11.1| |12.0| | | | | | | |


Expand Down
8 changes: 5 additions & 3 deletions src/CUDA2HIP_SPARSE_API_functions.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -863,7 +863,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_SPARSE_FUNCTION_MAP {
{"cusparseSpMV", {"hipsparseSpMV", "rocsparse_spmv", CONV_LIB_FUNC, API_SPARSE, 15}},
{"cusparseSpMV_bufferSize", {"hipsparseSpMV_bufferSize", "", CONV_LIB_FUNC, API_SPARSE, 15, ROC_UNSUPPORTED}},

{"cusparseSparseToDense", {"hipsparseSparseToDense", "", CONV_LIB_FUNC, API_SPARSE, 15, ROC_UNSUPPORTED}},
{"cusparseSparseToDense", {"hipsparseSparseToDense", "rocsparse_sparse_to_dense", CONV_LIB_FUNC, API_SPARSE, 15}},
{"cusparseSparseToDense_bufferSize", {"hipsparseSparseToDense_bufferSize", "", CONV_LIB_FUNC, API_SPARSE, 15, ROC_UNSUPPORTED}},
{"cusparseDenseToSparse_bufferSize", {"hipsparseDenseToSparse_bufferSize", "", CONV_LIB_FUNC, API_SPARSE, 15, ROC_UNSUPPORTED}},
{"cusparseDenseToSparse_analysis", {"hipsparseDenseToSparse_analysis", "", CONV_LIB_FUNC, API_SPARSE, 15, ROC_UNSUPPORTED}},
Expand Down Expand Up @@ -1223,7 +1223,7 @@ const std::map<llvm::StringRef, cudaAPIversions> CUDA_SPARSE_FUNCTION_VER_MAP {
{"cusparseCscSetPointers", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseCooSetPointers", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseSparseToDense_bufferSize", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseSparseToDense", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseSparseToDense", {CUDA_111, CUDA_0, CUDA_0 }},// A: CUSPARSE_VERSION 11300 C: CUSPARSE_VERSION 12000
{"cusparseDenseToSparse_bufferSize", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseDenseToSparse_analysis", {CUDA_111, CUDA_0, CUDA_0 }},
{"cusparseDenseToSparse_convert", {CUDA_111, CUDA_0, CUDA_0 }},
Expand Down Expand Up @@ -2423,6 +2423,7 @@ const std::map<llvm::StringRef, hipAPIversions> HIP_SPARSE_FUNCTION_VER_MAP {
{"rocsparse_cbsrilu0_buffer_size", {HIP_3080, HIP_0, HIP_0 }},
{"rocsparse_zbsrilu0_buffer_size", {HIP_3080, HIP_0, HIP_0 }},
{"rocsparse_csr2csc_buffer_size", {HIP_1090, HIP_0, HIP_0 }},
{"rocsparse_sparse_to_dense", {HIP_4010, HIP_0, HIP_0 }},
};

const std::map<llvm::StringRef, cudaAPIChangedVersions> CUDA_SPARSE_FUNCTION_CHANGED_VER_MAP {
Expand Down Expand Up @@ -2463,7 +2464,7 @@ const std::map<llvm::StringRef, cudaAPIChangedVersions> CUDA_SPARSE_FUNCTION_CHA
{"cusparseAxpby", {CUDA_120}},
{"cusparseGather", {CUDA_120}},
{"cusparseScatter", {CUDA_120}},
{"cusparseSparseToDense", {CUDA_120}},
{"cusparseSparseToDense", {CUDA_120}}, // C: CUSPARSE_VERSION 12000
{"cusparseSparseToDense_bufferSize", {CUDA_120}},
{"cusparseDenseToSparse_analysis", {CUDA_120}},
{"cusparseDenseToSparse_bufferSize", {CUDA_120}},
Expand Down Expand Up @@ -2525,6 +2526,7 @@ const std::map<llvm::StringRef, hipAPIChangedVersions> HIP_SPARSE_FUNCTION_CHANG
{"rocsparse_destroy_dnvec_descr", {HIP_6000}},
{"rocsparse_destroy_dnmat_descr", {HIP_6000}},
{"rocsparse_dnmat_get_strided_batch", {HIP_6000}},
{"rocsparse_sparse_to_dense", {HIP_6000}},
};

const std::map<unsigned int, llvm::StringRef> CUDA_SPARSE_API_SECTION_MAP {
Expand Down
13 changes: 12 additions & 1 deletion src/HipifyAction.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -204,6 +204,7 @@ const std::string sCusparseCbsrilu02_bufferSize = "cusparseCbsrilu02_bufferSize"
const std::string sCusparseDbsrilu02_bufferSize = "cusparseDbsrilu02_bufferSize";
const std::string sCusparseSbsrilu02_bufferSize = "cusparseSbsrilu02_bufferSize";
const std::string sCusparseCsr2cscEx2_bufferSize = "cusparseCsr2cscEx2_bufferSize";
const std::string sCusparseSparseToDense = "cusparseSparseToDense";

// CUDA_OVERLOADED
const std::string sCudaEventCreate = "cudaEventCreate";
Expand Down Expand Up @@ -1631,6 +1632,15 @@ std::map<std::string, ArgCastStruct> FuncArgCasts {
false
}
},
{sCusparseSparseToDense,
{
{
{4, {e_add_const_argument, cw_None, "nullptr"}}
},
true,
false
}
},
};

void HipifyAction::RewriteString(StringRef s, clang::SourceLocation start) {
Expand Down Expand Up @@ -2475,7 +2485,8 @@ std::unique_ptr<clang::ASTConsumer> HipifyAction::CreateASTConsumer(clang::Compi
sCusparseCbsrilu02_bufferSize,
sCusparseDbsrilu02_bufferSize,
sCusparseSbsrilu02_bufferSize,
sCusparseCsr2cscEx2_bufferSize
sCusparseCsr2cscEx2_bufferSize,
sCusparseSparseToDense
)
)
)
Expand Down
1 change: 1 addition & 0 deletions tests/lit.cfg
Original file line number Diff line number Diff line change
Expand Up @@ -99,6 +99,7 @@ if config.cuda_version_major < 12:
config.excludes.append('headers_test_08_12000.cu')
config.excludes.append('headers_test_09_12000.cu')
config.excludes.append('runtime_functions_12000.cu')
config.excludes.append('cusparse2rocsparse_12000.cu')

if config.cuda_version_major >= 12:
config.excludes.append('headers_test_06.cu')
Expand Down
1 change: 0 additions & 1 deletion tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu
Original file line number Diff line number Diff line change
Expand Up @@ -2581,7 +2581,6 @@ int main() {
// CHECK: status_t = hipsparseSparseToDense_bufferSize(handle_t, spmatA, dnmatB, sparseToDenseAlg_t, &bufferSize);
status_t = cusparseSparseToDense_bufferSize(handle_t, spmatA, dnmatB, sparseToDenseAlg_t, &bufferSize);

// TODO: Mark as C-Changed in 12.0.0
// CUDA: cusparseStatus_t CUSPARSEAPI cusparseSparseToDense(cusparseHandle_t handle, cusparseSpMatDescr_t matA, cusparseDnMatDescr_t matB, cusparseSparseToDenseAlg_t alg, void* buffer);
// HIP: HIPSPARSE_EXPORT hipsparseStatus_t hipsparseSparseToDense(hipsparseHandle_t handle, hipsparseSpMatDescr_t matA, hipsparseDnMatDescr_t matB, hipsparseSparseToDenseAlg_t alg, void* externalBuffer);
// CHECK: status_t = hipsparseSparseToDense(handle_t, spmatA, dnmatB, sparseToDenseAlg_t, tempBuffer);
Expand Down
108 changes: 108 additions & 0 deletions tests/unit_tests/synthetic/libraries/cusparse2rocsparse_12000.cu
Original file line number Diff line number Diff line change
@@ -0,0 +1,108 @@
// RUN: %run_test hipify "%s" "%t" %hipify_args 4 --skip-excluded-preprocessor-conditional-blocks --experimental --roc --use-hip-data-types %clang_args -ferror-limit=500

// CHECK: #include <hip/hip_runtime.h>
#include <cuda_runtime.h>
// CHECK: #include "hip/hip_complex.h"
#include "cuComplex.h"
#include <stdio.h>
// CHECK: #include "rocsparse.h"
#include "cusparse.h"
// CHECK-NOT: #include "rocsparse.h"

int main() {
printf("18.1. cuSPARSE API to rocSPARSE API synthetic test\n");

// CHECK: rocsparse_status status_t;
cusparseStatus_t status_t;

// CHECK: _rocsparse_handle *handle = nullptr;
// CHECK-NEXT: rocsparse_handle handle_t;
cusparseContext *handle = nullptr;
cusparseHandle_t handle_t;

// CHECK: _rocsparse_mat_descr *matDescr = nullptr;
// CHECK-NEXT: rocsparse_mat_descr matDescr_t, matDescr_t_2, matDescr_A, matDescr_B, matDescr_C, matDescr_D;
cusparseMatDescr *matDescr = nullptr;
cusparseMatDescr_t matDescr_t, matDescr_t_2, matDescr_A, matDescr_B, matDescr_C, matDescr_D;

// CHECK: rocsparse_operation opA, opB, opX;
cusparseOperation_t opA, opB, opX;

// CHECK: rocsparse_solve_policy solvePolicy_t;
cusparseSolvePolicy_t solvePolicy_t;

int m = 0;
int n = 0;
int k = 0;
int innz = 0;
int nnza = 0;
int nnzb = 0;
int nnzc = 0;
int nnzd = 0;
int csrRowPtrA = 0;
int csrRowPtrB = 0;
int csrRowPtrC = 0;
int csrRowPtrD = 0;
int csrColIndA = 0;
int csrColIndB = 0;
int csrColIndC = 0;
int csrColIndD = 0;
int bufferSizeInBytes = 0;
double dA = 0.f;
double dB = 0.f;
double dAlpha = 0.f;
double dF = 0.f;
double dX = 0.f;
double dcsrSortedValA = 0.f;
double dcsrSortedValB = 0.f;
double dcsrSortedValC = 0.f;
double dcsrSortedValD = 0.f;
float fAlpha = 0.f;
float fA = 0.f;
float fB = 0.f;
float fF = 0.f;
float fX = 0.f;
float csrSortedValA = 0.f;
float csrSortedValB = 0.f;
float csrSortedValC = 0.f;
float csrSortedValD = 0.f;
void *pBuffer = nullptr;
void *tempBuffer = nullptr;

// TODO: should be rocsparse_double_complex
// TODO: add to TypeOverloads cuDoubleComplex -> rocsparse_double_complex under a new option --sparse
// CHECK: rocblas_double_complex dcomplex, dcomplexA, dcomplexAlpha, dcomplexB, dcomplexBeta, dcomplexC, dcomplexF, dcomplexX, dcomplexY, dComplexbsrSortedValA, dComplexbsrSortedValC, dComplexcsrSortedValA, dComplexcsrSortedValB, dComplexcsrSortedValC, dComplexcsrSortedValD, dcomplextol, dComplexbsrSortedVal, dComplexbscVal, dComplexcscSortedVal, dcomplexds, dcomplexdl, dcomplexd, dcomplexdu, dcomplexdw, dcomplexx, dcomplex_boost_val;
cuDoubleComplex dcomplex, dcomplexA, dcomplexAlpha, dcomplexB, dcomplexBeta, dcomplexC, dcomplexF, dcomplexX, dcomplexY, dComplexbsrSortedValA, dComplexbsrSortedValC, dComplexcsrSortedValA, dComplexcsrSortedValB, dComplexcsrSortedValC, dComplexcsrSortedValD, dcomplextol, dComplexbsrSortedVal, dComplexbscVal, dComplexcscSortedVal, dcomplexds, dcomplexdl, dcomplexd, dcomplexdu, dcomplexdw, dcomplexx, dcomplex_boost_val;

// TODO: should be rocsparse_double_complex
// TODO: add to TypeOverloads cuComplex -> rocsparse_float_complex under a new option --sparse
// CHECK: rocblas_float_complex complex, complexA, complexAlpha, complexB, complexBeta, complexC, complexF, complexX, complexY, complexbsrValA, complexbsrSortedValC, complexcsrSortedValA, complexcsrSortedValB, complexcsrSortedValC, complexcsrSortedValD, complextol, complexbsrSortedVal, complexbscVal, complexcscSortedVal, complexds, complexdl, complexd, complexdu, complexdw, complexx, complex_boost_val;
cuComplex complex, complexA, complexAlpha, complexB, complexBeta, complexC, complexF, complexX, complexY, complexbsrValA, complexbsrSortedValC, complexcsrSortedValA, complexcsrSortedValB, complexcsrSortedValC, complexcsrSortedValD, complextol, complexbsrSortedVal, complexbscVal, complexcscSortedVal, complexds, complexdl, complexd, complexdu, complexdw, complexx, complex_boost_val;

#if CUDA_VERSION >= 11010 && CUSPARSE_VERSION >= 11300
// CHECK: rocsparse_sparse_to_dense_alg sparseToDenseAlg_t;
cusparseSparseToDenseAlg_t sparseToDenseAlg_t;
#endif

#if (CUDA_VERSION >= 10010 && CUDA_VERSION < 11000 && !defined(_WIN32)) || CUDA_VERSION >= 11000
// CHECK: rocsparse_spmat_descr spMatDescr_t, spmatA, spmatB, spmatC;
cusparseSpMatDescr_t spMatDescr_t, spmatA, spmatB, spmatC;

// CHECK: rocsparse_dnmat_descr dnMatDescr_t, dnmatA, dnmatB, dnmatC;
cusparseDnMatDescr_t dnMatDescr_t, dnmatA, dnmatB, dnmatC;
#endif

#if CUDA_VERSION >= 12000
// CHECK: rocsparse_const_spmat_descr constSpMatDescr = nullptr;
// CHECK-NEXT: rocsparse_const_spmat_descr constSpMatDescrB = nullptr;
cusparseConstSpMatDescr_t constSpMatDescr = nullptr;
cusparseConstSpMatDescr_t constSpMatDescrB = nullptr;

// CUDA: cusparseStatus_t CUSPARSEAPI cusparseSparseToDense(cusparseHandle_t handle, cusparseConstSpMatDescr_t matA, cusparseDnMatDescr_t matB, cusparseSparseToDenseAlg_t alg, void* externalBuffer);
// ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_sparse_to_dense(rocsparse_handle handle, rocsparse_const_spmat_descr mat_A, rocsparse_dnmat_descr mat_B, rocsparse_sparse_to_dense_alg alg, size_t* buffer_size, void* temp_buffer);
// CHECK: status_t = rocsparse_sparse_to_dense(handle_t, constSpMatDescr, dnmatB, sparseToDenseAlg_t, nullptr, tempBuffer);
status_t = cusparseSparseToDense(handle_t, constSpMatDescr, dnmatB, sparseToDenseAlg_t, tempBuffer);
#endif

return 0;
}
Original file line number Diff line number Diff line change
Expand Up @@ -67,6 +67,7 @@ int main() {
float csrSortedValC = 0.f;
float csrSortedValD = 0.f;
void *pBuffer = nullptr;
void *tempBuffer = nullptr;

// TODO: should be rocsparse_double_complex
// TODO: add to TypeOverloads cuDoubleComplex -> rocsparse_double_complex under a new option --sparse
Expand All @@ -78,6 +79,26 @@ int main() {
// CHECK: rocblas_float_complex complex, complexA, complexAlpha, complexB, complexBeta, complexC, complexF, complexX, complexY, complexbsrValA, complexbsrSortedValC, complexcsrSortedValA, complexcsrSortedValB, complexcsrSortedValC, complexcsrSortedValD, complextol, complexbsrSortedVal, complexbscVal, complexcscSortedVal, complexds, complexdl, complexd, complexdu, complexdw, complexx, complex_boost_val;
cuComplex complex, complexA, complexAlpha, complexB, complexBeta, complexC, complexF, complexX, complexY, complexbsrValA, complexbsrSortedValC, complexcsrSortedValA, complexcsrSortedValB, complexcsrSortedValC, complexcsrSortedValD, complextol, complexbsrSortedVal, complexbscVal, complexcscSortedVal, complexds, complexdl, complexd, complexdu, complexdw, complexx, complex_boost_val;

#if CUDA_VERSION >= 11010 && CUSPARSE_VERSION >= 11300
// CHECK: rocsparse_sparse_to_dense_alg sparseToDenseAlg_t;
cusparseSparseToDenseAlg_t sparseToDenseAlg_t;
#endif

#if (CUDA_VERSION >= 10010 && CUDA_VERSION < 11000 && !defined(_WIN32)) || CUDA_VERSION >= 11000
// CHECK: rocsparse_spmat_descr spMatDescr_t, spmatA, spmatB, spmatC;
cusparseSpMatDescr_t spMatDescr_t, spmatA, spmatB, spmatC;

// CHECK: rocsparse_dnmat_descr dnMatDescr_t, dnmatA, dnmatB, dnmatC;
cusparseDnMatDescr_t dnMatDescr_t, dnmatA, dnmatB, dnmatC;

#if CUDA_VERSION < 12000
// CUDA: cusparseStatus_t CUSPARSEAPI cusparseSparseToDense(cusparseHandle_t handle, cusparseSpMatDescr_t matA, cusparseDnMatDescr_t matB, cusparseSparseToDenseAlg_t alg, void* buffer);
// ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_sparse_to_dense(rocsparse_handle handle, const rocsparse_spmat_descr mat_A, rocsparse_dnmat_descr mat_B, rocsparse_sparse_to_dense_alg alg, size_t* buffer_size, void* temp_buffer);
// CHECK: status_t = rocsparse_sparse_to_dense(handle_t, spmatA, dnmatB, sparseToDenseAlg_t, nullptr, tempBuffer);
status_t = cusparseSparseToDense(handle_t, spmatA, dnmatB, sparseToDenseAlg_t, tempBuffer);
#endif
#endif

#if CUDA_VERSION < 12000
// CHECK: rocsparse_mat_descr csrsv2_info;
csrsv2Info_t csrsv2_info;
Expand Down

0 comments on commit 5d82b98

Please sign in to comment.