From 472483dea5a8331ffdc4061fce1ba1aa1aae36eb Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Sun, 10 Sep 2023 19:32:51 +0200 Subject: [PATCH] [HIPIFY][#674][#1014][rocSPARSE][feature] rocSPARSE support - Step 39 - functions + Implemented a new function call transformation type "replace argument with a const" (`e_replace_argument_with_const`) + The new transformation is tested on rocSPARSE functions `rocsparse_(s|d|c|z)csrilu0`, where: - the penultimate argument `cusparseSolvePolicy_t policy` should always be the const value `rocsparse_solve_policy_auto`; - it is how `hipsparse(S|D|C|Z)csrilu02` calls `rocsparse_(s|d|c|z)csrilu0` in its implementation; + Updated synthetic tests and the regenerated hipify-perl and SPARSE docs + Added the missing `csrilu02Info_t` -> `rocsparse_mat_info` transformation --- bin/hipify-perl | 6 ++ .../CUSPARSE_API_supported_by_HIP_and_ROC.md | 12 ++-- docs/tables/CUSPARSE_API_supported_by_ROC.md | 12 ++-- src/CUDA2HIP_SPARSE_API_functions.cpp | 12 ++-- src/CUDA2HIP_SPARSE_API_types.cpp | 4 +- src/CUDA2HIP_Scripting.h | 3 +- src/HipifyAction.cpp | 59 ++++++++++++++++++- .../synthetic/libraries/cusparse2hipsparse.cu | 21 +++++++ .../synthetic/libraries/cusparse2rocsparse.cu | 23 ++++++++ 9 files changed, 130 insertions(+), 22 deletions(-) diff --git a/bin/hipify-perl b/bin/hipify-perl index 83ffc211..070fcff3 100755 --- a/bin/hipify-perl +++ b/bin/hipify-perl @@ -1682,6 +1682,7 @@ sub rocSubstitutions { subst("cusparseCcsr2gebsr_bufferSize", "rocsparse_ccsr2gebsr_buffer_size", "library"); subst("cusparseCcsr2hyb", "rocsparse_ccsr2hyb", "library"); subst("cusparseCcsrcolor", "rocsparse_ccsrcolor", "library"); + subst("cusparseCcsrilu02", "rocsparse_ccsrilu0", "library"); subst("cusparseCdense2csc", "rocsparse_cdense2csc", "library"); subst("cusparseCdense2csr", "rocsparse_cdense2csr", "library"); subst("cusparseCgebsr2csr", "rocsparse_cgebsr2csr", "library"); @@ -1732,6 +1733,7 @@ sub rocSubstitutions { subst("cusparseDcsr2gebsr_bufferSize", "rocsparse_dcsr2gebsr_buffer_size", "library"); subst("cusparseDcsr2hyb", "rocsparse_dcsr2hyb", "library"); subst("cusparseDcsrcolor", "rocsparse_dcsrcolor", "library"); + subst("cusparseDcsrilu02", "rocsparse_dcsrilu0", "library"); subst("cusparseDdense2csc", "rocsparse_ddense2csc", "library"); subst("cusparseDdense2csr", "rocsparse_ddense2csr", "library"); subst("cusparseDestroy", "rocsparse_destroy_handle", "library"); @@ -1801,6 +1803,7 @@ sub rocSubstitutions { subst("cusparseScsr2gebsr_bufferSize", "rocsparse_scsr2gebsr_buffer_size", "library"); subst("cusparseScsr2hyb", "rocsparse_scsr2hyb", "library"); subst("cusparseScsrcolor", "rocsparse_scsrcolor", "library"); + subst("cusparseScsrilu02", "rocsparse_scsrilu0", "library"); subst("cusparseSdense2csc", "rocsparse_sdense2csc", "library"); subst("cusparseSdense2csr", "rocsparse_sdense2csr", "library"); subst("cusparseSetMatDiagType", "rocsparse_set_mat_diag_type", "library"); @@ -1873,6 +1876,7 @@ sub rocSubstitutions { subst("cusparseZcsr2gebsr_bufferSize", "rocsparse_zcsr2gebsr_buffer_size", "library"); subst("cusparseZcsr2hyb", "rocsparse_zcsr2hyb", "library"); subst("cusparseZcsrcolor", "rocsparse_zcsrcolor", "library"); + subst("cusparseZcsrilu02", "rocsparse_zcsrilu0", "library"); subst("cusparseZdense2csc", "rocsparse_zdense2csc", "library"); subst("cusparseZdense2csr", "rocsparse_zdense2csr", "library"); subst("cusparseZgebsr2csr", "rocsparse_zgebsr2csr", "library"); @@ -1894,6 +1898,8 @@ sub rocSubstitutions { subst("cusparseZnnz_compress", "rocsparse_znnz_compress", "library"); subst("cublas.h", "rocblas.h", "include_cuda_main_header"); subst("cublas_v2.h", "rocblas.h", "include_cuda_main_header_v2"); + subst("csrilu02Info", "_rocsparse_mat_info", "type"); + subst("csrilu02Info_t", "rocsparse_mat_info", "type"); subst("cuComplex", "rocblas_float_complex", "type"); subst("cuDoubleComplex", "rocblas_double_complex", "type"); subst("cuFloatComplex", "rocblas_float_complex", "type"); 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 a6520195..bc9f4e90 100644 --- a/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md +++ b/docs/tables/CUSPARSE_API_supported_by_HIP_and_ROC.md @@ -120,8 +120,8 @@ |`csrgemm2Info_t`| | |12.0|`csrgemm2Info_t`|2.8.0| | | | | | | | | |`csric02Info`| |12.2| |`csric02Info`|3.1.0| | | | | | | | | |`csric02Info_t`| |12.2| |`csric02Info_t`|3.1.0| | | | | | | | | -|`csrilu02Info`| |12.2| |`csrilu02Info`|1.9.2| | | | | | | | | -|`csrilu02Info_t`| |12.2| |`csrilu02Info_t`|1.9.2| | | | | | | | | +|`csrilu02Info`| |12.2| |`csrilu02Info`|1.9.2| | | |`_rocsparse_mat_info`|1.9.0| | | | +|`csrilu02Info_t`| |12.2| |`csrilu02Info_t`|1.9.2| | | |`rocsparse_mat_info`|1.9.0| | | | |`csrsm2Info`|9.2| |12.0| | | | | | | | | | | |`csrsm2Info_t`|9.2| |12.0|`csrsm2Info_t`|3.1.0| | | | | | | | | |`csrsv2Info`| | |12.0| | | | | | | | | | | @@ -484,7 +484,7 @@ |`cusparseCcsric02_bufferSize`| |12.2| |`hipsparseCcsric02_bufferSize`|3.1.0| | | | | | | | | |`cusparseCcsric02_bufferSizeExt`| |12.2| |`hipsparseCcsric02_bufferSizeExt`|3.1.0| | | | | | | | | |`cusparseCcsrilu0`| |10.2|11.0| | | | | | | | | | | -|`cusparseCcsrilu02`| |12.2| |`hipsparseCcsrilu02`|3.1.0| | | | | | | | | +|`cusparseCcsrilu02`| |12.2| |`hipsparseCcsrilu02`|3.1.0| | | |`rocsparse_ccsrilu0`|2.10.0| | | | |`cusparseCcsrilu02_analysis`| |12.2| |`hipsparseCcsrilu02_analysis`|3.1.0| | | | | | | | | |`cusparseCcsrilu02_bufferSize`| |12.2| |`hipsparseCcsrilu02_bufferSize`|3.1.0| | | | | | | | | |`cusparseCcsrilu02_bufferSizeExt`| |12.2| |`hipsparseCcsrilu02_bufferSizeExt`|3.1.0| | | | | | | | | @@ -518,7 +518,7 @@ |`cusparseDcsric02_bufferSize`| |12.2| |`hipsparseDcsric02_bufferSize`|3.1.0| | | | | | | | | |`cusparseDcsric02_bufferSizeExt`| |12.2| |`hipsparseDcsric02_bufferSizeExt`|3.1.0| | | | | | | | | |`cusparseDcsrilu0`| |10.2|11.0| | | | | | | | | | | -|`cusparseDcsrilu02`| |12.2| |`hipsparseDcsrilu02`|1.9.2| | | | | | | | | +|`cusparseDcsrilu02`| |12.2| |`hipsparseDcsrilu02`|1.9.2| | | |`rocsparse_dcsrilu0`|1.9.0| | | | |`cusparseDcsrilu02_analysis`| |12.2| |`hipsparseDcsrilu02_analysis`|1.9.2| | | | | | | | | |`cusparseDcsrilu02_bufferSize`| |12.2| |`hipsparseDcsrilu02_bufferSize`|1.9.2| | | | | | | | | |`cusparseDcsrilu02_bufferSizeExt`| |12.2| |`hipsparseDcsrilu02_bufferSizeExt`|1.9.2| | | | | | | | | @@ -551,7 +551,7 @@ |`cusparseScsric02_bufferSize`| |12.2| |`hipsparseScsric02_bufferSize`|3.1.0| | | | | | | | | |`cusparseScsric02_bufferSizeExt`| |12.2| |`hipsparseScsric02_bufferSizeExt`|3.1.0| | | | | | | | | |`cusparseScsrilu0`| |10.2|11.0| | | | | | | | | | | -|`cusparseScsrilu02`| |12.2| |`hipsparseScsrilu02`|1.9.2| | | | | | | | | +|`cusparseScsrilu02`| |12.2| |`hipsparseScsrilu02`|1.9.2| | | |`rocsparse_scsrilu0`|1.9.0| | | | |`cusparseScsrilu02_analysis`| |12.2| |`hipsparseScsrilu02_analysis`|1.9.2| | | | | | | | | |`cusparseScsrilu02_bufferSize`| |12.2| |`hipsparseScsrilu02_bufferSize`|1.9.2| | | | | | | | | |`cusparseScsrilu02_bufferSizeExt`| |12.2| |`hipsparseScsrilu02_bufferSizeExt`|1.9.2| | | | | | | | | @@ -588,7 +588,7 @@ |`cusparseZcsric02_bufferSize`| |12.2| |`hipsparseZcsric02_bufferSize`|3.1.0| | | | | | | | | |`cusparseZcsric02_bufferSizeExt`| |12.2| |`hipsparseZcsric02_bufferSizeExt`|3.1.0| | | | | | | | | |`cusparseZcsrilu0`| |10.2|11.0| | | | | | | | | | | -|`cusparseZcsrilu02`| |12.2| |`hipsparseZcsrilu02`|3.1.0| | | | | | | | | +|`cusparseZcsrilu02`| |12.2| |`hipsparseZcsrilu02`|3.1.0| | | |`rocsparse_zcsrilu0`|2.10.0| | | | |`cusparseZcsrilu02_analysis`| |12.2| |`hipsparseZcsrilu02_analysis`|3.1.0| | | | | | | | | |`cusparseZcsrilu02_bufferSize`| |12.2| |`hipsparseZcsrilu02_bufferSize`|3.1.0| | | | | | | | | |`cusparseZcsrilu02_bufferSizeExt`| |12.2| |`hipsparseZcsrilu02_bufferSizeExt`|3.1.0| | | | | | | | | diff --git a/docs/tables/CUSPARSE_API_supported_by_ROC.md b/docs/tables/CUSPARSE_API_supported_by_ROC.md index c2dad2f2..880e08df 100644 --- a/docs/tables/CUSPARSE_API_supported_by_ROC.md +++ b/docs/tables/CUSPARSE_API_supported_by_ROC.md @@ -120,8 +120,8 @@ |`csrgemm2Info_t`| | |12.0| | | | | | |`csric02Info`| |12.2| | | | | | | |`csric02Info_t`| |12.2| | | | | | | -|`csrilu02Info`| |12.2| | | | | | | -|`csrilu02Info_t`| |12.2| | | | | | | +|`csrilu02Info`| |12.2| |`_rocsparse_mat_info`|1.9.0| | | | +|`csrilu02Info_t`| |12.2| |`rocsparse_mat_info`|1.9.0| | | | |`csrsm2Info`|9.2| |12.0| | | | | | |`csrsm2Info_t`|9.2| |12.0| | | | | | |`csrsv2Info`| | |12.0| | | | | | @@ -484,7 +484,7 @@ |`cusparseCcsric02_bufferSize`| |12.2| | | | | | | |`cusparseCcsric02_bufferSizeExt`| |12.2| | | | | | | |`cusparseCcsrilu0`| |10.2|11.0| | | | | | -|`cusparseCcsrilu02`| |12.2| | | | | | | +|`cusparseCcsrilu02`| |12.2| |`rocsparse_ccsrilu0`|2.10.0| | | | |`cusparseCcsrilu02_analysis`| |12.2| | | | | | | |`cusparseCcsrilu02_bufferSize`| |12.2| | | | | | | |`cusparseCcsrilu02_bufferSizeExt`| |12.2| | | | | | | @@ -518,7 +518,7 @@ |`cusparseDcsric02_bufferSize`| |12.2| | | | | | | |`cusparseDcsric02_bufferSizeExt`| |12.2| | | | | | | |`cusparseDcsrilu0`| |10.2|11.0| | | | | | -|`cusparseDcsrilu02`| |12.2| | | | | | | +|`cusparseDcsrilu02`| |12.2| |`rocsparse_dcsrilu0`|1.9.0| | | | |`cusparseDcsrilu02_analysis`| |12.2| | | | | | | |`cusparseDcsrilu02_bufferSize`| |12.2| | | | | | | |`cusparseDcsrilu02_bufferSizeExt`| |12.2| | | | | | | @@ -551,7 +551,7 @@ |`cusparseScsric02_bufferSize`| |12.2| | | | | | | |`cusparseScsric02_bufferSizeExt`| |12.2| | | | | | | |`cusparseScsrilu0`| |10.2|11.0| | | | | | -|`cusparseScsrilu02`| |12.2| | | | | | | +|`cusparseScsrilu02`| |12.2| |`rocsparse_scsrilu0`|1.9.0| | | | |`cusparseScsrilu02_analysis`| |12.2| | | | | | | |`cusparseScsrilu02_bufferSize`| |12.2| | | | | | | |`cusparseScsrilu02_bufferSizeExt`| |12.2| | | | | | | @@ -588,7 +588,7 @@ |`cusparseZcsric02_bufferSize`| |12.2| | | | | | | |`cusparseZcsric02_bufferSizeExt`| |12.2| | | | | | | |`cusparseZcsrilu0`| |10.2|11.0| | | | | | -|`cusparseZcsrilu02`| |12.2| | | | | | | +|`cusparseZcsrilu02`| |12.2| |`rocsparse_zcsrilu0`|2.10.0| | | | |`cusparseZcsrilu02_analysis`| |12.2| | | | | | | |`cusparseZcsrilu02_bufferSize`| |12.2| | | | | | | |`cusparseZcsrilu02_bufferSizeExt`| |12.2| | | | | | | diff --git a/src/CUDA2HIP_SPARSE_API_functions.cpp b/src/CUDA2HIP_SPARSE_API_functions.cpp index 8b8a1462..341641fd 100644 --- a/src/CUDA2HIP_SPARSE_API_functions.cpp +++ b/src/CUDA2HIP_SPARSE_API_functions.cpp @@ -391,10 +391,10 @@ const std::map CUDA_SPARSE_FUNCTION_MAP { {"cusparseCcsrilu02_analysis", {"hipsparseCcsrilu02_analysis", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, {"cusparseZcsrilu02_analysis", {"hipsparseZcsrilu02_analysis", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"cusparseScsrilu02", {"hipsparseScsrilu02", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"cusparseDcsrilu02", {"hipsparseDcsrilu02", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"cusparseCcsrilu02", {"hipsparseCcsrilu02", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"cusparseZcsrilu02", {"hipsparseZcsrilu02", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, + {"cusparseScsrilu02", {"hipsparseScsrilu02", "rocsparse_scsrilu0", CONV_LIB_FUNC, API_SPARSE, 12, CUDA_DEPRECATED}}, + {"cusparseDcsrilu02", {"hipsparseDcsrilu02", "rocsparse_dcsrilu0", CONV_LIB_FUNC, API_SPARSE, 12, CUDA_DEPRECATED}}, + {"cusparseCcsrilu02", {"hipsparseCcsrilu02", "rocsparse_ccsrilu0", CONV_LIB_FUNC, API_SPARSE, 12, CUDA_DEPRECATED}}, + {"cusparseZcsrilu02", {"hipsparseZcsrilu02", "rocsparse_zcsrilu0", CONV_LIB_FUNC, API_SPARSE, 12, CUDA_DEPRECATED}}, {"cusparseXcsrilu02_zeroPivot", {"hipsparseXcsrilu02_zeroPivot", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, {"cusparseSbsrilu02_numericBoost", {"hipsparseSbsrilu02_numericBoost", "", CONV_LIB_FUNC, API_SPARSE, 12, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, @@ -2175,6 +2175,10 @@ const std::map HIP_SPARSE_FUNCTION_VER_MAP { {"rocsparse_cgtsv_buffer_size", {HIP_4030, HIP_0, HIP_0 }}, {"rocsparse_dgtsv_buffer_size", {HIP_4030, HIP_0, HIP_0 }}, {"rocsparse_sgtsv_buffer_size", {HIP_4030, HIP_0, HIP_0 }}, + {"rocsparse_zcsrilu0", {HIP_2100, HIP_0, HIP_0 }}, + {"rocsparse_ccsrilu0", {HIP_2100, HIP_0, HIP_0 }}, + {"rocsparse_dcsrilu0", {HIP_1090, HIP_0, HIP_0 }}, + {"rocsparse_scsrilu0", {HIP_1090, HIP_0, HIP_0 }}, }; const std::map CUDA_SPARSE_API_SECTION_MAP { diff --git a/src/CUDA2HIP_SPARSE_API_types.cpp b/src/CUDA2HIP_SPARSE_API_types.cpp index 3650d3b1..8c53510a 100644 --- a/src/CUDA2HIP_SPARSE_API_types.cpp +++ b/src/CUDA2HIP_SPARSE_API_types.cpp @@ -53,8 +53,8 @@ const std::map CUDA_SPARSE_TYPE_NAME_MAP { {"bsric02Info", {"bsric02Info", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, {"bsric02Info_t", {"bsric02Info_t", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"csrilu02Info", {"csrilu02Info", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, - {"csrilu02Info_t", {"csrilu02Info_t", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, + {"csrilu02Info", {"csrilu02Info", "_rocsparse_mat_info", CONV_TYPE, API_SPARSE, 4, CUDA_DEPRECATED}}, + {"csrilu02Info_t", {"csrilu02Info_t", "rocsparse_mat_info", CONV_TYPE, API_SPARSE, 4, CUDA_DEPRECATED}}, {"bsrilu02Info", {"bsrilu02Info", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, {"bsrilu02Info_t", {"bsrilu02Info_t", "", CONV_TYPE, API_SPARSE, 4, ROC_UNSUPPORTED | CUDA_DEPRECATED}}, diff --git a/src/CUDA2HIP_Scripting.h b/src/CUDA2HIP_Scripting.h index 9f689a94..6fb68cef 100644 --- a/src/CUDA2HIP_Scripting.h +++ b/src/CUDA2HIP_Scripting.h @@ -33,6 +33,7 @@ namespace hipify { e_add_const_argument, e_add_var_argument, e_move_argument, + e_replace_argument_with_const, }; enum OverloadTypes { @@ -51,7 +52,7 @@ namespace hipify { struct CastInfo { CastTypes castType; CastWarning castWarn; - std::string constValToAdd = ""; + std::string constValToAddOrReplace = ""; unsigned moveOrCopyTo = 0; unsigned numberToMoveOrCopy = 1; }; diff --git a/src/HipifyAction.cpp b/src/HipifyAction.cpp index e61703ad..f2e02501 100644 --- a/src/HipifyAction.cpp +++ b/src/HipifyAction.cpp @@ -98,6 +98,10 @@ const std::string sCusparseZgtsvInterleavedBatch_bufferSizeExt = "cusparseZgtsvI const std::string sCusparseCgtsvInterleavedBatch_bufferSizeExt = "cusparseCgtsvInterleavedBatch_bufferSizeExt"; const std::string sCusparseDgtsvInterleavedBatch_bufferSizeExt = "cusparseDgtsvInterleavedBatch_bufferSizeExt"; const std::string sCusparseSgtsvInterleavedBatch_bufferSizeExt = "cusparseSgtsvInterleavedBatch_bufferSizeExt"; +const std::string sCusparseZcsrilu02 = "cusparseZcsrilu02"; +const std::string sCusparseCcsrilu02 = "cusparseCcsrilu02"; +const std::string sCusparseDcsrilu02 = "cusparseDcsrilu02"; +const std::string sCusparseScsrilu02 = "cusparseScsrilu02"; // CUDA_OVERLOADED const std::string sCudaEventCreate = "cudaEventCreate"; const std::string sCudaGraphInstantiate = "cudaGraphInstantiate"; @@ -122,6 +126,7 @@ std::string getCastType(hipify::CastTypes c) { case e_add_const_argument: return ""; case e_add_var_argument: return ""; case e_move_argument: return ""; + case e_replace_argument_with_const: return ""; default: return ""; } } @@ -542,6 +547,42 @@ std::map FuncArgCasts { false } }, + {sCusparseZcsrilu02, + { + { + {8, {e_replace_argument_with_const, cw_None, "rocsparse_solve_policy_auto"}} + }, + true, + false + } + }, + {sCusparseCcsrilu02, + { + { + {8, {e_replace_argument_with_const, cw_None, "rocsparse_solve_policy_auto"}} + }, + true, + false + } + }, + {sCusparseDcsrilu02, + { + { + {8, {e_replace_argument_with_const, cw_None, "rocsparse_solve_policy_auto"}} + }, + true, + false + } + }, + {sCusparseScsrilu02, + { + { + {8, {e_replace_argument_with_const, cw_None, "rocsparse_solve_policy_auto"}} + }, + true, + false + } + }, }; void HipifyAction::RewriteString(StringRef s, clang::SourceLocation start) { @@ -1048,9 +1089,9 @@ bool HipifyAction::cudaHostFuncCall(const mat::MatchFinder::MatchResult &Result) case e_add_const_argument: { if (argNum < call->getNumArgs()) - OS << c.second.constValToAdd << ", "; + OS << c.second.constValToAddOrReplace << ", "; else - OS << ", " << c.second.constValToAdd; + OS << ", " << c.second.constValToAddOrReplace; break; } case e_add_var_argument: @@ -1070,6 +1111,14 @@ bool HipifyAction::cudaHostFuncCall(const mat::MatchFinder::MatchResult &Result) } break; } + case e_replace_argument_with_const: + { + if (argNum >= call->getNumArgs()) + break; + OS << c.second.constValToAddOrReplace; + length = SM->getCharacterData(clang::Lexer::getLocForEndOfToken(e, 0, *SM, DefaultLangOptions)) - SM->getCharacterData(s); + break; + } default: OS << getCastType(c.second.castType) << "(" << readSourceText(*SM, sr) << ")"; length = SM->getCharacterData(clang::Lexer::getLocForEndOfToken(e, 0, *SM, DefaultLangOptions)) - SM->getCharacterData(s); @@ -1259,7 +1308,11 @@ std::unique_ptr HipifyAction::CreateASTConsumer(clang::Compi sCusparseZgtsvInterleavedBatch_bufferSizeExt, sCusparseCgtsvInterleavedBatch_bufferSizeExt, sCusparseDgtsvInterleavedBatch_bufferSizeExt, - sCusparseSgtsvInterleavedBatch_bufferSizeExt + sCusparseSgtsvInterleavedBatch_bufferSizeExt, + sCusparseZcsrilu02, + sCusparseCcsrilu02, + sCusparseDcsrilu02, + sCusparseScsrilu02 ) ) ) diff --git a/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu b/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu index ef869489..14901333 100644 --- a/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu +++ b/tests/unit_tests/synthetic/libraries/cusparse2hipsparse.cu @@ -248,6 +248,7 @@ int main() { float fdw = 0.f; float fx = 0.f; pruneInfo_t prune_info; + csrilu02Info_t csrilu02_info; // CHECK: hipDoubleComplex dcomplex, dcomplexA, dcomplexB, dComplexbsrSortedValA, dComplexbsrSortedValC, dComplexcsrSortedValA, dComplexcsrSortedValC, dcomplextol, dComplexbsrSortedVal, dComplexbscVal, dComplexcscSortedVal, dcomplexds, dcomplexdl, dcomplexd, dcomplexdu, dcomplexdw, dcomplexx; cuDoubleComplex dcomplex, dcomplexA, dcomplexB, dComplexbsrSortedValA, dComplexbsrSortedValC, dComplexcsrSortedValA, dComplexcsrSortedValC, dcomplextol, dComplexbsrSortedVal, dComplexbscVal, dComplexcscSortedVal, dcomplexds, dcomplexdl, dcomplexd, dcomplexdu, dcomplexdw, dcomplexx; @@ -678,6 +679,26 @@ int main() { // CHECK: status_t = hipsparseSgtsv2StridedBatch_bufferSizeExt(handle_t, m, &fdl, &fd, &fdu, &fx, batchCount, ibatchStride, &bufferSize); status_t = cusparseSgtsv2StridedBatch_bufferSizeExt(handle_t, m, &fdl, &fd, &fdu, &fx, batchCount, ibatchStride, &bufferSize); + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseZcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, cuDoubleComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // HIP: HIPSPARSE_EXPORT hipsparseStatus_t hipsparseZcsrilu02(hipsparseHandle_t handle, int m, int nnz, const hipsparseMatDescr_t descrA, hipDoubleComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, hipsparseSolvePolicy_t policy, void* pBuffer); + // CHECK: status_t = hipsparseZcsrilu02(handle_t, m, innz, matDescr_A, &dComplexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + status_t = cusparseZcsrilu02(handle_t, m, innz, matDescr_A, &dComplexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseCcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, cuComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // HIP: HIPSPARSE_EXPORT hipsparseStatus_t hipsparseCcsrilu02(hipsparseHandle_t handle, int m, int nnz, const hipsparseMatDescr_t descrA, hipComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, hipsparseSolvePolicy_t policy, void* pBuffer); + // CHECK: status_t = hipsparseCcsrilu02(handle_t, m, innz, matDescr_A, &complexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + status_t = cusparseCcsrilu02(handle_t, m, innz, matDescr_A, &complexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseDcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, double* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // HIP: HIPSPARSE_EXPORT hipsparseStatus_t hipsparseDcsrilu02(hipsparseHandle_t handle, int m, int nnz, const hipsparseMatDescr_t descrA, double* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, hipsparseSolvePolicy_t policy, void* pBuffer); + // CHECK: status_t = hipsparseDcsrilu02(handle_t, m, innz, matDescr_A, &dcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + status_t = cusparseDcsrilu02(handle_t, m, innz, matDescr_A, &dcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseScsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, float* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // HIP: HIPSPARSE_EXPORT hipsparseStatus_t hipsparseScsrilu02(hipsparseHandle_t handle, int m, int nnz, const hipsparseMatDescr_t descrA, float* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, hipsparseSolvePolicy_t policy, void* pBuffer); + // CHECK: status_t = hipsparseScsrilu02(handle_t, m, innz, matDescr_A, &csrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + status_t = cusparseScsrilu02(handle_t, m, innz, matDescr_A, &csrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + #if CUDA_VERSION >= 8000 // CHECK: hipDataType dataType_t; // CHECK-NEXT: hipDataType dataType; diff --git a/tests/unit_tests/synthetic/libraries/cusparse2rocsparse.cu b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse.cu index 64e72e4a..bcf5617f 100644 --- a/tests/unit_tests/synthetic/libraries/cusparse2rocsparse.cu +++ b/tests/unit_tests/synthetic/libraries/cusparse2rocsparse.cu @@ -251,6 +251,9 @@ int main() { // CHECK: rocsparse_mat_info prune_info; pruneInfo_t prune_info; + // CHECK: rocsparse_mat_info csrilu02_info; + csrilu02Info_t csrilu02_info; + // 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, dcomplexB, dComplexbsrSortedValA, dComplexbsrSortedValC, dComplexcsrSortedValA, dComplexcsrSortedValC, dcomplextol, dComplexbsrSortedVal, dComplexbscVal, dComplexcscSortedVal, dcomplexds, dcomplexdl, dcomplexd, dcomplexdu, dcomplexdw, dcomplexx; @@ -684,6 +687,26 @@ int main() { // CHECK: status_t = rocsparse_sgtsv_no_pivot_strided_batch_buffer_size(handle_t, m, &fdl, &fd, &fdu, &fx, batchCount, ibatchStride, &bufferSize); status_t = cusparseSgtsv2StridedBatch_bufferSizeExt(handle_t, m, &fdl, &fd, &fdu, &fx, batchCount, ibatchStride, &bufferSize); + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseZcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, cuDoubleComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_zcsrilu0(rocsparse_handle handle, rocsparse_int m, rocsparse_int nnz, const rocsparse_mat_descr descr, rocsparse_double_complex* csr_val, const rocsparse_int* csr_row_ptr, const rocsparse_int* csr_col_ind, rocsparse_mat_info info, rocsparse_solve_policy policy, void* temp_buffer); + // CHECK: status_t = rocsparse_zcsrilu0(handle_t, m, innz, matDescr_A, &dComplexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, rocsparse_solve_policy_auto, pBuffer); + status_t = cusparseZcsrilu02(handle_t, m, innz, matDescr_A, &dComplexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseCcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, cuComplex* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_ccsrilu0(rocsparse_handle handle, rocsparse_int m, rocsparse_int nnz, const rocsparse_mat_descr descr, rocsparse_float_complex* csr_val, const rocsparse_int* csr_row_ptr, const rocsparse_int* csr_col_ind, rocsparse_mat_info info, rocsparse_solve_policy policy, void* temp_buffer); + // CHECK: status_t = rocsparse_ccsrilu0(handle_t, m, innz, matDescr_A, &complexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, rocsparse_solve_policy_auto, pBuffer); + status_t = cusparseCcsrilu02(handle_t, m, innz, matDescr_A, &complexcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseDcsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, double* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_dcsrilu0(rocsparse_handle handle, rocsparse_int m, rocsparse_int nnz, const rocsparse_mat_descr descr, double* csr_val, const rocsparse_int* csr_row_ptr, const rocsparse_int* csr_col_ind, rocsparse_mat_info info, rocsparse_solve_policy policy, void* temp_buffer); + // CHECK: status_t = rocsparse_dcsrilu0(handle_t, m, innz, matDescr_A, &dcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, rocsparse_solve_policy_auto, pBuffer); + status_t = cusparseDcsrilu02(handle_t, m, innz, matDescr_A, &dcsrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + + // CUDA: CUSPARSE_DEPRECATED cusparseStatus_t CUSPARSEAPI cusparseScsrilu02(cusparseHandle_t handle, int m, int nnz, const cusparseMatDescr_t descrA, float* csrSortedValA_valM, const int* csrSortedRowPtrA, const int* csrSortedColIndA, csrilu02Info_t info, cusparseSolvePolicy_t policy, void* pBuffer); + // ROC: ROCSPARSE_EXPORT rocsparse_status rocsparse_scsrilu0(rocsparse_handle handle, rocsparse_int m, rocsparse_int nnz, const rocsparse_mat_descr descr, float* csr_val, const rocsparse_int* csr_row_ptr, const rocsparse_int* csr_col_ind, rocsparse_mat_info info, rocsparse_solve_policy policy, void* temp_buffer); + // CHECK: status_t = rocsparse_scsrilu0(handle_t, m, innz, matDescr_A, &csrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, rocsparse_solve_policy_auto, pBuffer); + status_t = cusparseScsrilu02(handle_t, m, innz, matDescr_A, &csrSortedValA, &csrRowPtrA, &csrColIndA, csrilu02_info, solvePolicy_t, pBuffer); + #if CUDA_VERSION >= 8000 // CHECK: hipDataType dataType_t; // TODO: [#899] There should be rocsparse_datatype