From b74240c7f7e61940018f3556181c237cad854b9a Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Fri, 31 Mar 2023 20:34:50 +0200 Subject: [PATCH] [HIPIFY][#584][DNN][MIOpen] cuDNN -> MIOpen - Part 7 + Continued supporting hipification to MIOpen based on `miopen.h` + Updated the synthetic test `cudnn2miopen.cu` accordingly [ToDo] + [feature] Arguments swap + [feature] Additional argument (const value, firstly) --- src/CUDA2HIP_DNN_API_functions.cpp | 6 +-- src/CUDA2HIP_DNN_API_types.cpp | 14 +++---- .../synthetic/libraries/cudnn2miopen.cu | 41 +++++++++++++++++++ 3 files changed, 51 insertions(+), 10 deletions(-) diff --git a/src/CUDA2HIP_DNN_API_functions.cpp b/src/CUDA2HIP_DNN_API_functions.cpp index 31b246b0..8bb8ae9f 100644 --- a/src/CUDA2HIP_DNN_API_functions.cpp +++ b/src/CUDA2HIP_DNN_API_functions.cpp @@ -111,11 +111,11 @@ const std::map CUDA_DNN_FUNCTION_MAP { {"cudnnDestroyConvolutionDescriptor", {"hipdnnDestroyConvolutionDescriptor", "miopenDestroyConvolutionDescriptor", CONV_LIB_FUNC, API_DNN, 2}}, {"cudnnGetConvolutionForwardAlgorithmMaxCount", {"hipdnnGetConvolutionForwardAlgorithmMaxCount", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}}, {"cudnnFindConvolutionForwardAlgorithm", {"hipdnnFindConvolutionForwardAlgorithm", "", CONV_LIB_FUNC, API_DNN, 2}}, - {"cudnnFindConvolutionForwardAlgorithmEx", {"hipdnnFindConvolutionForwardAlgorithmEx", "", CONV_LIB_FUNC, API_DNN, 2}}, + {"cudnnFindConvolutionForwardAlgorithmEx", {"hipdnnFindConvolutionForwardAlgorithmEx", "miopenFindConvolutionForwardAlgorithm", CONV_LIB_FUNC, API_DNN, 2}}, {"cudnnGetConvolutionForwardAlgorithm", {"hipdnnGetConvolutionForwardAlgorithm", "", CONV_LIB_FUNC, API_DNN, 2, CUDA_DEPRECATED | CUDA_REMOVED}}, {"cudnnGetConvolutionForwardAlgorithm_v7", {"hipdnnGetConvolutionForwardAlgorithm_v7", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}}, - {"cudnnGetConvolutionForwardWorkspaceSize", {"hipdnnGetConvolutionForwardWorkspaceSize", "", CONV_LIB_FUNC, API_DNN, 2}}, - {"cudnnConvolutionForward", {"hipdnnConvolutionForward", "", CONV_LIB_FUNC, API_DNN, 2}}, + {"cudnnGetConvolutionForwardWorkspaceSize", {"hipdnnGetConvolutionForwardWorkspaceSize", "miopenConvolutionForwardGetWorkSpaceSize", CONV_LIB_FUNC, API_DNN, 2}}, + {"cudnnConvolutionForward", {"hipdnnConvolutionForward", "miopenConvolutionForward", CONV_LIB_FUNC, API_DNN, 2}}, {"cudnnConvolutionBiasActivationForward", {"hipdnnConvolutionBiasActivationForward", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}}, {"cudnnConvolutionBackwardBias", {"hipdnnConvolutionBackwardBias", "", CONV_LIB_FUNC, API_DNN, 2}}, {"cudnnGetConvolutionBackwardFilterAlgorithmMaxCount", {"hipdnnGetConvolutionBackwardFilterAlgorithmMaxCount", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}}, diff --git a/src/CUDA2HIP_DNN_API_types.cpp b/src/CUDA2HIP_DNN_API_types.cpp index 0f303b5b..db43b054 100644 --- a/src/CUDA2HIP_DNN_API_types.cpp +++ b/src/CUDA2HIP_DNN_API_types.cpp @@ -149,12 +149,12 @@ const std::map CUDA_DNN_TYPE_NAME_MAP { {"cudnnNanPropagation_t", {"hipdnnNanPropagation_t", "miopenNanPropagation_t", CONV_TYPE, API_DNN, 1}}, {"CUDNN_NOT_PROPAGATE_NAN", {"HIPDNN_NOT_PROPAGATE_NAN", "MIOPEN_NOT_PROPAGATE_NAN", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 0 {"CUDNN_PROPAGATE_NAN", {"HIPDNN_PROPAGATE_NAN", "MIOPEN_PROPAGATE_NAN", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 1 - {"cudnnConvolutionBwdDataAlgo_t", {"hipdnnConvolutionBwdDataAlgo_t", "", CONV_TYPE, API_DNN, 1}}, - {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_0", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_0", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 0 - {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_1", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_1", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 1 - {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_FFT", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_FFT", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 2 + {"cudnnConvolutionBwdDataAlgo_t", {"hipdnnConvolutionBwdDataAlgo_t", "miopenConvBwdDataAlgorithm_t", CONV_TYPE, API_DNN, 1}}, + {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_0", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_0", "miopenConvolutionBwdDataAlgoGEMM", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 0 + {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_1", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_1", "miopenConvolutionBwdDataAlgoDirect", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 1 + {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_FFT", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_FFT", "miopenConvolutionBwdDataAlgoFFT", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 2 {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_FFT_TILING", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_FFT_TILING", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 3 - {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 4 + {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD", "miopenConvolutionBwdDataAlgoWinograd", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 4 {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD_NONFUSED", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD_NONFUSED", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 5 {"CUDNN_CONVOLUTION_BWD_DATA_ALGO_COUNT", {"HIPDNN_CONVOLUTION_BWD_DATA_ALGO_TRANSPOSE_GEMM", "", CONV_NUMERIC_LITERAL, API_DNN, 1}}, // 6 {"cudnnConvolutionBwdFilterAlgo_t", {"hipdnnConvolutionBwdFilterAlgo_t", "", CONV_TYPE, API_DNN, 1}}, @@ -825,8 +825,8 @@ const std::map CUDA_DNN_TYPE_NAME_MAP { {"cudnnCTCLossDescriptor_t", {"hipdnnCTCLossDescriptor_t", "miopenCTCLossDescriptor_t", CONV_TYPE, API_DNN, 1, HIP_UNSUPPORTED}}, {"cudnnTensorTransformStruct", {"hipdnnTensorTransformStruct", "", CONV_TYPE, API_DNN, 1, HIP_UNSUPPORTED}}, {"cudnnTensorTransformDescriptor_t", {"hipdnnTensorTransformDescriptor_t", "", CONV_TYPE, API_DNN, 1, HIP_UNSUPPORTED}}, - {"cudnnConvolutionFwdAlgoPerf_t", {"hipdnnConvolutionFwdAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}}, - {"cudnnConvolutionFwdAlgoPerfStruct", {"hipdnnConvolutionFwdAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}}, + {"cudnnConvolutionFwdAlgoPerf_t", {"hipdnnConvolutionFwdAlgoPerf_t", "miopenConvAlgoPerf_t", CONV_TYPE, API_DNN, 1}}, + {"cudnnConvolutionFwdAlgoPerfStruct", {"hipdnnConvolutionFwdAlgoPerf_t", "miopenConvAlgoPerf_t", CONV_TYPE, API_DNN, 1}}, {"cudnnConvolutionBwdFilterAlgoPerf_t", {"hipdnnConvolutionBwdFilterAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}}, {"cudnnConvolutionBwdFilterAlgoPerfStruct", {"hipdnnConvolutionBwdFilterAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}}, {"cudnnConvolutionBwdDataAlgoPerf_t", {"hipdnnConvolutionBwdDataAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}}, diff --git a/tests/unit_tests/synthetic/libraries/cudnn2miopen.cu b/tests/unit_tests/synthetic/libraries/cudnn2miopen.cu index 8235dd1a..7cec72f8 100644 --- a/tests/unit_tests/synthetic/libraries/cudnn2miopen.cu +++ b/tests/unit_tests/synthetic/libraries/cudnn2miopen.cu @@ -218,6 +218,22 @@ int main() { cudnnIndicesType_t _16BIT_INDICES = CUDNN_16BIT_INDICES; cudnnIndicesType_t _8BIT_INDICES = CUDNN_8BIT_INDICES; + // CHECK: miopenConvBwdDataAlgorithm_t ConvolutionBwdDataAlgo_t; + // CHECK-NEXT: miopenConvBwdDataAlgorithm_t CONVOLUTION_BWD_DATA_ALGO_0 = miopenConvolutionBwdDataAlgoGEMM; + // CHECK-NEXT: miopenConvBwdDataAlgorithm_t CONVOLUTION_BWD_DATA_ALGO_1 = miopenConvolutionBwdDataAlgoDirect; + // CHECK-NEXT: miopenConvBwdDataAlgorithm_t CONVOLUTION_BWD_DATA_ALGO_FFT = miopenConvolutionBwdDataAlgoFFT; + // CHECK-NEXT: miopenConvBwdDataAlgorithm_t CONVOLUTION_BWD_DATA_ALGO_WINOGRAD = miopenConvolutionBwdDataAlgoWinograd; + cudnnConvolutionBwdDataAlgo_t ConvolutionBwdDataAlgo_t; + cudnnConvolutionBwdDataAlgo_t CONVOLUTION_BWD_DATA_ALGO_0 = CUDNN_CONVOLUTION_BWD_DATA_ALGO_0; + cudnnConvolutionBwdDataAlgo_t CONVOLUTION_BWD_DATA_ALGO_1 = CUDNN_CONVOLUTION_BWD_DATA_ALGO_1; + cudnnConvolutionBwdDataAlgo_t CONVOLUTION_BWD_DATA_ALGO_FFT = CUDNN_CONVOLUTION_BWD_DATA_ALGO_FFT; + cudnnConvolutionBwdDataAlgo_t CONVOLUTION_BWD_DATA_ALGO_WINOGRAD = CUDNN_CONVOLUTION_BWD_DATA_ALGO_WINOGRAD; + + // CHECK: miopenConvAlgoPerf_t ConvolutionFwdAlgoPerf_t; + // CHECK-NEXT: miopenConvAlgoPerf_t ConvolutionFwdAlgoPerfStruct; + cudnnConvolutionFwdAlgoPerf_t ConvolutionFwdAlgoPerf_t; + cudnnConvolutionFwdAlgoPerfStruct ConvolutionFwdAlgoPerfStruct; + // CUDA: cudnnStatus_t CUDNNWINAPI cudnnCreateTensorDescriptor(cudnnTensorDescriptor_t* tensorDesc); // MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenCreateTensorDescriptor(miopenTensorDescriptor_t* tensorDesc); // CHECK: status = miopenCreateTensorDescriptor(&tensorDescriptor); @@ -256,6 +272,7 @@ int main() { cudnnTensorDescriptor_t cD; cudnnTensorDescriptor_t xD; cudnnTensorDescriptor_t yD; + cudnnTensorDescriptor_t wD; cudnnTensorDescriptor_t inputD; void* A = nullptr; void* B = nullptr; @@ -266,7 +283,12 @@ int main() { void* beta = nullptr; void* x = nullptr; void* y = nullptr; + void* W = nullptr; int groupCount = 0; + int requestedAlgoCount = 0; + int returnedAlgoCount = 0; + void* workSpace = nullptr; + size_t workSpaceSizeInBytes = 0; // TODO: cudnnOpTensor -> miopenOpTensor: different signatures: cudnnOpTensorDescriptor_t != miopenTensorOp_t // CUDA: cudnnStatus_t CUDNNWINAPI cudnnOpTensor(cudnnHandle_t handle, const cudnnOpTensorDescriptor_t opTensorDesc, const void* alpha1, const cudnnTensorDescriptor_t aDesc, const void* A, const void* alpha2, const cudnnTensorDescriptor_t bDesc, const void* B, const void* beta, const cudnnTensorDescriptor_t cDesc, void* C); @@ -307,5 +329,24 @@ int main() { // CHECK: status = miopenDestroyConvolutionDescriptor(convolutionDescriptor); status = cudnnDestroyConvolutionDescriptor(convolutionDescriptor); + // TODO: add additional bool argument = true + // CUDA: cudnnStatus_t CUDNNWINAPI cudnnFindConvolutionForwardAlgorithmEx(cudnnHandle_t handle, const cudnnTensorDescriptor_t xDesc, const void* x, const cudnnFilterDescriptor_t wDesc, const void* w, const cudnnConvolutionDescriptor_t convDesc, const cudnnTensorDescriptor_t yDesc, void* y, const int requestedAlgoCount, int* returnedAlgoCount, cudnnConvolutionFwdAlgoPerf_t* perfResults, void* workSpace, size_t workSpaceSizeInBytes); + // MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenFindConvolutionForwardAlgorithm(miopenHandle_t handle, const miopenTensorDescriptor_t xDesc, const void* x, const miopenTensorDescriptor_t wDesc, const void* w, const miopenConvolutionDescriptor_t convDesc, const miopenTensorDescriptor_t yDesc, void* y, const int requestAlgoCount, int* returnedAlgoCount, miopenConvAlgoPerf_t* perfResults, void* workSpace, size_t workSpaceSize, bool exhaustiveSearch); + // CHECK: status = miopenFindConvolutionForwardAlgorithm(handle, xD, x, filterDescriptor, W, convolutionDescriptor, yD, y, requestedAlgoCount, &returnedAlgoCount, &ConvolutionFwdAlgoPerf_t, workSpace, workSpaceSizeInBytes); + status = cudnnFindConvolutionForwardAlgorithmEx(handle, xD, x, filterDescriptor, W, convolutionDescriptor, yD, y, requestedAlgoCount, &returnedAlgoCount, &ConvolutionFwdAlgoPerf_t, workSpace, workSpaceSizeInBytes); + + // TODO: remove the penultimate arg (cudnnConvolutionFwdAlgo_t algo) + // TODO: swap 2 and 3 arguments (const miopenTensorDescriptor_t wDesc and const miopenTensorDescriptor_t xDesc) + // CUDA: cudnnStatus_t CUDNNWINAPI cudnnGetConvolutionForwardWorkspaceSize(cudnnHandle_t handle, const cudnnTensorDescriptor_t xDesc, const cudnnFilterDescriptor_t wDesc, const cudnnConvolutionDescriptor_t convDesc, const cudnnTensorDescriptor_t yDesc, cudnnConvolutionFwdAlgo_t algo, size_t* sizeInBytes); + // MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenConvolutionForwardGetWorkSpaceSize(miopenHandle_t handle, const miopenTensorDescriptor_t wDesc, const miopenTensorDescriptor_t xDesc, const miopenConvolutionDescriptor_t convDesc, const miopenTensorDescriptor_t yDesc, size_t* workSpaceSize); + // CHECK: status = miopenConvolutionForwardGetWorkSpaceSize(handle, xD, filterDescriptor, convolutionDescriptor, yD, convolutionFwdAlgo , &workSpaceSizeInBytes); + status = cudnnGetConvolutionForwardWorkspaceSize(handle, xD, filterDescriptor, convolutionDescriptor, yD, convolutionFwdAlgo , &workSpaceSizeInBytes); + + // TODO: swap correstly last 5 arguments + // CUDA: cudnnStatus_t CUDNNWINAPI cudnnConvolutionForward(cudnnHandle_t handle, const void* alpha, const cudnnTensorDescriptor_t xDesc, const void* x, const cudnnFilterDescriptor_t wDesc, const void* w, const cudnnConvolutionDescriptor_t convDesc, cudnnConvolutionFwdAlgo_t algo, void* workSpace, size_t workSpaceSizeInBytes, const void* beta, const cudnnTensorDescriptor_t yDesc, void* y); + // MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenConvolutionForward(miopenHandle_t handle, const void* alpha, const miopenTensorDescriptor_t xDesc, const void* x, const miopenTensorDescriptor_t wDesc, const void* w, const miopenConvolutionDescriptor_t convDesc, miopenConvFwdAlgorithm_t algo, const void* beta, const miopenTensorDescriptor_t yDesc, void* y, void* workSpace, size_t workSpaceSize); + // CHECK: status = miopenConvolutionForward(handle, alpha, xD, x, filterDescriptor, W, convolutionDescriptor, convolutionFwdAlgo, workSpace, workSpaceSizeInBytes, beta, yD, y); + status = cudnnConvolutionForward(handle, alpha, xD, x, filterDescriptor, W, convolutionDescriptor, convolutionFwdAlgo, workSpace, workSpaceSizeInBytes, beta, yD, y); + return 0; }