Skip to content

Commit

Permalink
Merge pull request #823 from emankov/HIPIFY
Browse files Browse the repository at this point in the history 
[HIPIFY][#584][DNN][MIOpen] cuDNN -> MIOpen - Part 9
  • Loading branch information
@emankov
emankov authored Apr 1, 2023
2 parents 31d10de + fabfc2e commit 82edd28
Show file tree
Hide file tree
Showing 4 changed files with 59 additions and 8 deletions.
8 changes: 4 additions & 4 deletions src/CUDA2HIP_DNN_API_functions.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -117,7 +117,7 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DNN_FUNCTION_MAP {
{"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}},
{"cudnnConvolutionBackwardBias", {"hipdnnConvolutionBackwardBias", "miopenConvolutionBackwardBias", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnGetConvolutionBackwardFilterAlgorithmMaxCount", {"hipdnnGetConvolutionBackwardFilterAlgorithmMaxCount", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}},
{"cudnnFindConvolutionBackwardFilterAlgorithm", {"hipdnnFindConvolutionBackwardFilterAlgorithm", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnFindConvolutionBackwardFilterAlgorithmEx", {"hipdnnFindConvolutionBackwardFilterAlgorithmEx", "", CONV_LIB_FUNC, API_DNN, 2}},
Expand All @@ -130,15 +130,15 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DNN_FUNCTION_MAP {
{"cudnnFindConvolutionBackwardDataAlgorithmEx", {"hipdnnFindConvolutionBackwardDataAlgorithmEx", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnGetConvolutionBackwardDataAlgorithm", {"hipdnnGetConvolutionBackwardDataAlgorithm", "", CONV_LIB_FUNC, API_DNN, 2, CUDA_DEPRECATED | CUDA_REMOVED}},
{"cudnnGetConvolutionBackwardDataAlgorithm_v7", {"hipdnnGetConvolutionBackwardDataAlgorithm_v7", "", CONV_LIB_FUNC, API_DNN, 2, HIP_UNSUPPORTED}},
{"cudnnGetConvolutionBackwardDataWorkspaceSize", {"hipdnnGetConvolutionBackwardDataWorkspaceSize", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnConvolutionBackwardData", {"hipdnnConvolutionBackwardData", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnGetConvolutionBackwardDataWorkspaceSize", {"hipdnnGetConvolutionBackwardDataWorkspaceSize", "miopenConvolutionBackwardDataGetWorkSpaceSize", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnConvolutionBackwardData", {"hipdnnConvolutionBackwardData", "miopenConvolutionBackwardData", CONV_LIB_FUNC, API_DNN, 2}},

// cuDNN Sortmax functions
{"cudnnSoftmaxForward", {"hipdnnSoftmaxForward", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnSoftmaxBackward", {"hipdnnSoftmaxBackward", "", CONV_LIB_FUNC, API_DNN, 2}},

// cuDNN Pooling functions
{"cudnnCreatePoolingDescriptor", {"hipdnnCreatePoolingDescriptor", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnCreatePoolingDescriptor", {"hipdnnCreatePoolingDescriptor", "miopenCreatePoolingDescriptor", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnSetPooling2dDescriptor", {"hipdnnSetPooling2dDescriptor", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnGetPooling2dDescriptor", {"hipdnnGetPooling2dDescriptor", "", CONV_LIB_FUNC, API_DNN, 2}},
{"cudnnSetPoolingNdDescriptor", {"hipdnnSetPoolingNdDescriptor", "", CONV_LIB_FUNC, API_DNN, 2}},
Expand Down
4 changes: 2 additions & 2 deletions src/CUDA2HIP_DNN_API_types.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -829,8 +829,8 @@ const std::map<llvm::StringRef, hipCounter> CUDA_DNN_TYPE_NAME_MAP {
{"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}},
{"cudnnConvolutionBwdDataAlgoPerfStruct", {"hipdnnConvolutionBwdDataAlgoPerf_t", "", CONV_TYPE, API_DNN, 1}},
{"cudnnConvolutionBwdDataAlgoPerf_t", {"hipdnnConvolutionBwdDataAlgoPerf_t", "miopenConvAlgoPerf_t", CONV_TYPE, API_DNN, 1}},
{"cudnnConvolutionBwdDataAlgoPerfStruct", {"hipdnnConvolutionBwdDataAlgoPerf_t", "miopenConvAlgoPerf_t", CONV_TYPE, API_DNN, 1}},
{"cudnnDropoutStruct", {"hipdnnDropoutStruct", "", CONV_TYPE, API_DNN, 1, HIP_UNSUPPORTED}},
{"cudnnDropoutDescriptor_t", {"hipdnnDropoutDescriptor_t", "miopenDropoutDescriptor_t", CONV_TYPE, API_DNN, 1}},
{"cudnnAlgorithmStruct", {"hipdnnAlgorithmStruct", "", CONV_TYPE, API_DNN, 1, HIP_UNSUPPORTED}},
Expand Down
13 changes: 12 additions & 1 deletion src/HipifyAction.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -68,6 +68,7 @@ const std::string sCuOccupancyMaxPotentialBlockSize = "cuOccupancyMaxPotentialBl
const std::string sCuOccupancyMaxPotentialBlockSizeWithFlags = "cuOccupancyMaxPotentialBlockSizeWithFlags";
const std::string sCudaGetTextureReference = "cudaGetTextureReference";
const std::string sCudnnGetConvolutionForwardWorkspaceSize = "cudnnGetConvolutionForwardWorkspaceSize";
const std::string sCudnnGetConvolutionBackwardDataWorkspaceSize = "cudnnGetConvolutionBackwardDataWorkspaceSize";
// Matchers' names
const StringRef sCudaLaunchKernel = "cudaLaunchKernel";
const StringRef sCudaHostFuncCall = "cudaHostFuncCall";
Expand Down Expand Up @@ -203,6 +204,15 @@ std::map<std::string, ArgCastStruct> FuncArgCasts {
true
}
},
{sCudnnGetConvolutionBackwardDataWorkspaceSize,
{
{
{5, {e_remove_argument, cw_None}}
},
true,
true
}
},
};

void HipifyAction::RewriteString(StringRef s, clang::SourceLocation start) {
Expand Down Expand Up @@ -764,7 +774,8 @@ std::unique_ptr<clang::ASTConsumer> HipifyAction::CreateASTConsumer(clang::Compi
sCuOccupancyMaxPotentialBlockSize,
sCuOccupancyMaxPotentialBlockSizeWithFlags,
sCudaGetTextureReference,
sCudnnGetConvolutionForwardWorkspaceSize
sCudnnGetConvolutionForwardWorkspaceSize,
sCudnnGetConvolutionBackwardDataWorkspaceSize
)
)
)
Expand Down
42 changes: 41 additions & 1 deletion tests/unit_tests/synthetic/libraries/cudnn2miopen.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -234,6 +234,11 @@ int main() {
cudnnConvolutionFwdAlgoPerf_t ConvolutionFwdAlgoPerf_t;
cudnnConvolutionFwdAlgoPerfStruct ConvolutionFwdAlgoPerfStruct;

// CHECK: miopenConvAlgoPerf_t ConvolutionBwdDataAlgoPerf_t;
// CHECK-NEXT: miopenConvAlgoPerf_t ConvolutionBwdDataAlgoPerfStruct;
cudnnConvolutionBwdDataAlgoPerf_t ConvolutionBwdDataAlgoPerf_t;
cudnnConvolutionBwdDataAlgoPerfStruct ConvolutionBwdDataAlgoPerfStruct;

// CUDA: cudnnStatus_t CUDNNWINAPI cudnnCreateTensorDescriptor(cudnnTensorDescriptor_t* tensorDesc);
// MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenCreateTensorDescriptor(miopenTensorDescriptor_t* tensorDesc);
// CHECK: status = miopenCreateTensorDescriptor(&tensorDescriptor);
Expand Down Expand Up @@ -267,13 +272,23 @@ int main() {
// CHECK: status = miopenDestroyTensorDescriptor(tensorDescriptor);
status = cudnnDestroyTensorDescriptor(tensorDescriptor);


// CHECK: miopenTensorDescriptor_t aD;
// CHECK-NEXT: miopenTensorDescriptor_t bD;
// CHECK-NEXT: miopenTensorDescriptor_t cD;
// CHECK-NEXT: miopenTensorDescriptor_t xD;
// CHECK-NEXT: miopenTensorDescriptor_t yD;
// CHECK-NEXT: miopenTensorDescriptor_t wD;
// CHECK-NEXT: miopenTensorDescriptor_t inputD;
// CHECK-NEXT: miopenTensorDescriptor_t dbD;
cudnnTensorDescriptor_t aD;
cudnnTensorDescriptor_t bD;
cudnnTensorDescriptor_t cD;
cudnnTensorDescriptor_t xD;
cudnnTensorDescriptor_t yD;
cudnnTensorDescriptor_t wD;
cudnnTensorDescriptor_t inputD;
cudnnTensorDescriptor_t dbD;
void* A = nullptr;
void* B = nullptr;
void* C = nullptr;
Expand All @@ -282,8 +297,11 @@ int main() {
void* alpha2 = nullptr;
void* beta = nullptr;
void* x = nullptr;
void* dx = nullptr;
void* y = nullptr;
void* dy = nullptr;
void* W = nullptr;
void* db = nullptr;
int groupCount = 0;
int requestedAlgoCount = 0;
int returnedAlgoCount = 0;
Expand Down Expand Up @@ -341,11 +359,33 @@ int main() {
// CHECK: status = miopenConvolutionForwardGetWorkSpaceSize(handle, xD, filterDescriptor, convolutionDescriptor, yD, &workSpaceSizeInBytes);
status = cudnnGetConvolutionForwardWorkspaceSize(handle, xD, filterDescriptor, convolutionDescriptor, yD, convolutionFwdAlgo, &workSpaceSizeInBytes);

// TODO: swap correstly last 5 arguments
// TODO: swap correctly 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);

// TODO: swap 2 and 3 arguments
// CUDA: cudnnStatus_t CUDNNWINAPI cudnnGetConvolutionBackwardDataWorkspaceSize(cudnnHandle_t handle, const cudnnFilterDescriptor_t wDesc, const cudnnTensorDescriptor_t dyDesc, const cudnnConvolutionDescriptor_t convDesc, const cudnnTensorDescriptor_t dxDesc, cudnnConvolutionBwdDataAlgo_t algo, size_t* sizeInBytes);
// MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenConvolutionBackwardDataGetWorkSpaceSize(miopenHandle_t handle, const miopenTensorDescriptor_t dyDesc, const miopenTensorDescriptor_t wDesc, const miopenConvolutionDescriptor_t convDesc, const miopenTensorDescriptor_t dxDesc, size_t* workSpaceSize);
// CHECK: status = miopenConvolutionBackwardDataGetWorkSpaceSize(handle, filterDescriptor, yD, convolutionDescriptor, xD, &workSpaceSizeInBytes);
status = cudnnGetConvolutionBackwardDataWorkspaceSize(handle, filterDescriptor, yD, convolutionDescriptor, xD, ConvolutionBwdDataAlgo_t, &workSpaceSizeInBytes);

// TODO: swap correctly all args, starting from 3rd
// CUDA: cudnnStatus_t CUDNNWINAPI cudnnConvolutionBackwardData(cudnnHandle_t handle, const void* alpha, const cudnnFilterDescriptor_t wDesc, const void* w, const cudnnTensorDescriptor_t dyDesc, const void* dy, const cudnnConvolutionDescriptor_t convDesc, cudnnConvolutionBwdDataAlgo_t algo, void* workSpace, size_t workSpaceSizeInBytes, const void* beta, const cudnnTensorDescriptor_t dxDesc, void* dx);
// MIOPEN MIOPEN_EXPORT miopenStatus_t miopenConvolutionBackwardData(miopenHandle_t handle, const void* alpha, const miopenTensorDescriptor_t dyDesc, const void* dy, const miopenTensorDescriptor_t wDesc, const void* w, const miopenConvolutionDescriptor_t convDesc, miopenConvBwdDataAlgorithm_t algo, const void* beta, const miopenTensorDescriptor_t dxDesc, void* dx, void* workSpace, size_t workSpaceSize);
// CHECK: status = miopenConvolutionBackwardData(handle, alpha, filterDescriptor, W, yD, dy, convolutionDescriptor, ConvolutionBwdDataAlgo_t, workSpace, workSpaceSizeInBytes, beta, xD, dx);
status = cudnnConvolutionBackwardData(handle, alpha, filterDescriptor, W, yD, dy, convolutionDescriptor, ConvolutionBwdDataAlgo_t, workSpace, workSpaceSizeInBytes, beta, xD, dx);

// CUDA: cudnnStatus_t CUDNNWINAPI cudnnConvolutionBackwardBias(cudnnHandle_t handle, const void* alpha, const cudnnTensorDescriptor_t dyDesc, const void* dy, const void* beta, const cudnnTensorDescriptor_t dbDesc, void* db);
// MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenConvolutionBackwardBias(miopenHandle_t handle, const void* alpha, const miopenTensorDescriptor_t dyDesc, const void* dy, const void* beta, const miopenTensorDescriptor_t dbDesc, void* db);
// CHECK: status = miopenConvolutionBackwardBias(handle, alpha, yD, dy, beta, dbD, db);
status = cudnnConvolutionBackwardBias(handle, alpha, yD, dy, beta, dbD, db);

// CUDA: cudnnStatus_t CUDNNWINAPI cudnnCreatePoolingDescriptor(cudnnPoolingDescriptor_t* poolingDesc);
// MIOPEN: MIOPEN_EXPORT miopenStatus_t miopenCreatePoolingDescriptor(miopenPoolingDescriptor_t* poolDesc);
// CHECK: status = miopenCreatePoolingDescriptor(&poolingDescriptor);
status = cudnnCreatePoolingDescriptor(&poolingDescriptor);

return 0;
}

0 comments on commit 82edd28

Please sign in to comment.