From 4c282f4f251a4fc7e5855ee637f6c661033bb02d Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Tue, 16 Jan 2024 14:43:39 +0000 Subject: [PATCH] [HIPIFY][#674][rocSPARSE][feature] rocSPARSE support - Step 98 - function `rocsparse_sparse_to_dense` + [IMP] `rocsparse_sparse_to_dense` has been changed in 6.0.0; so reflected that in HIPIFY, docs, and tests + Updated `SPARSE` synthetic tests, the regenerated hipify-perl, and `SPARSE` `CUDA2HIP` documentation --- bin/hipify-perl | 1 + .../CUSPARSE_API_supported_by_HIP_and_ROC.md | 2 +- docs/tables/CUSPARSE_API_supported_by_ROC.md | 2 +- src/CUDA2HIP_SPARSE_API_functions.cpp | 8 +- src/HipifyAction.cpp | 13 ++- tests/lit.cfg | 1 + .../synthetic/libraries/cusparse2hipsparse.cu | 1 - .../libraries/cusparse2rocsparse_12000.cu | 108 ++++++++++++++++++ .../cusparse2rocsparse_before_12000.cu | 21 ++++ 9 files changed, 150 insertions(+), 7 deletions(-) create mode 100644 tests/unit_tests/synthetic/libraries/cusparse2rocsparse_12000.cu diff --git a/bin/hipify-perl b/bin/hipify-perl index da2969a7..bb822efb 100755 --- a/bin/hipify-perl +++ b/bin/hipify-perl @@ -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"); diff --git a/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md b/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md index 231cca71..d613af43 100644 --- a/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md +++ b/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md @@ -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| | | | | | | | | diff --git a/docs/tables/CUSPARSE_API_supported_by_ROC.md b/docs/tables/CUSPARSE_API_supported_by_ROC.md index b2ee9621..8bc35bc7 100644 --- a/docs/tables/CUSPARSE_API_supported_by_ROC.md +++ b/docs/tables/CUSPARSE_API_supported_by_ROC.md @@ -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| | | | | | | | diff --git a/src/CUDA2HIP_SPARSE_API_functions.cpp b/src/CUDA2HIP_SPARSE_API_functions.cpp index 144886bc..0acf3237 100644 --- a/src/CUDA2HIP_SPARSE_API_functions.cpp +++ b/src/CUDA2HIP_SPARSE_API_functions.cpp @@ -863,7 +863,7 @@ const std::map 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}}, @@ -1223,7 +1223,7 @@ const std::map 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 }}, @@ -2423,6 +2423,7 @@ const std::map 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 CUDA_SPARSE_FUNCTION_CHANGED_VER_MAP { @@ -2463,7 +2464,7 @@ const std::map 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}}, @@ -2525,6 +2526,7 @@ const std::map 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 CUDA_SPARSE_API_SECTION_MAP { diff --git a/src/HipifyAction.cpp b/src/HipifyAction.cpp index 33dae471..23ecf5c3 100644 --- a/src/HipifyAction.cpp +++ b/src/HipifyAction.cpp @@ -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"; @@ -1631,6 +1632,15 @@ std::map FuncArgCasts { false } }, + {sCusparseSparseToDense, + { + { + {4, {e_add_const_argument, cw_None, "nullptr"}} + }, + true, + false + } + }, }; void HipifyAction::RewriteString(StringRef s, clang::SourceLocation start) { @@ -2475,7 +2485,8 @@ std::unique_ptr HipifyAction::CreateASTConsumer(clang::Compi sCusparseCbsrilu02_bufferSize, sCusparseDbsrilu02_bufferSize, sCusparseSbsrilu02_bufferSize, - sCusparseCsr2cscEx2_bufferSize + sCusparseCsr2cscEx2_bufferSize, + sCusparseSparseToDense ) ) ) diff --git a/tests/lit.cfg b/tests/lit.cfg index 26f47a13..1075b807 100644 --- a/tests/lit.cfg +++ b/tests/lit.cfg @@ -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') diff --git a/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu b/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu index c5d41291..9639cc65 100644 --- a/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu +++ b/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu @@ -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); diff --git a/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_12000.cu b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_12000.cu new file mode 100644 index 00000000..1d27eaae --- /dev/null +++ b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_12000.cu @@ -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 +#include +// CHECK: #include "hip/hip_complex.h" +#include "cuComplex.h" +#include +// 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; +} diff --git a/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_before_12000.cu b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_before_12000.cu index 68796a9f..636c4828 100644 --- a/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_before_12000.cu +++ b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse_before_12000.cu @@ -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 @@ -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;