From 3a57fc0a4b63fca77d7eff7aaa3574e2b9493693 Mon Sep 17 00:00:00 2001 From: Evgeny Mankov Date: Sat, 23 Feb 2019 20:46:22 +0300 Subject: [PATCH] [HIPIFY][Caffe2] Initial Caffe2 support [ROCm/hip commit: 2619f22e5ca71d602a70f405feb8918708ad0cc2] --- projects/hip/hipify-clang/src/CUDA2HIP.cpp | 13 +++ projects/hip/hipify-clang/src/CUDA2HIP.h | 4 + .../src/CUDA2HIP_CAFFE2_API_functions.cpp | 28 +++++ .../src/CUDA2HIP_CAFFE2_API_types.cpp | 34 ++++++ .../src/CUDA2HIP_Complex_API_functions.cpp | 2 +- .../src/CUDA2HIP_SPARSE_API_functions.cpp | 2 +- projects/hip/hipify-clang/src/Statistics.cpp | 3 +- projects/hip/hipify-clang/src/Statistics.h | 1 + projects/hip/tests/hipify-clang/lit.cfg | 2 + .../CAFFE2/caffe2/core/common_cudnn.h | 7 ++ .../caffe2/operators/spatial_batch_norm_op.h | 14 +++ .../unit_tests/libraries/CAFFE2/caffe2_01.cu | 12 ++ .../unit_tests/libraries/CAFFE2/caffe2_02.cu | 103 ++++++++++++++++++ 13 files changed, 222 insertions(+), 3 deletions(-) create mode 100644 projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_functions.cpp create mode 100644 projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_types.cpp create mode 100644 projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/core/common_cudnn.h create mode 100644 projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/operators/spatial_batch_norm_op.h create mode 100644 projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_01.cu create mode 100644 projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_02.cu diff --git a/projects/hip/hipify-clang/src/CUDA2HIP.cpp b/projects/hip/hipify-clang/src/CUDA2HIP.cpp index 7536c5c0f7..b9e97e874d 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP.cpp @@ -64,6 +64,17 @@ const std::map CUDA_INCLUDE_MAP{ // cuBLAS includes {"cusparse.h", {"hipsparse.h", "", CONV_INCLUDE_CUDA_MAIN_H, API_SPARSE}}, {"cusparse_v2.h", {"hipsparse.h", "", CONV_INCLUDE_CUDA_MAIN_H, API_SPARSE}}, + // CAFFE2 includes + {"caffe2/core/common_gpu.h", {"caffe2/core/hip/common_gpu.h", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/core/context_gpu.h", {"caffe2/core/hip/context_gpu.h", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/operators/operator_fallback_gpu.h", {"", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/operators/spatial_batch_norm_op.h", {"caffe2/operators/hip/spatial_batch_norm_op_miopen.hip", "", CONV_INCLUDE, API_CAFFE2}}, + {"caffe2/operators/generate_proposals_op_util_nms_gpu.h", {"", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/operators/max_pool_with_index_gpu.h", {"", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/operators/rnn/recurrent_network_executor_gpu.h", {"", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/utils/math/reduce.cuh", {"caffe2/utils/math/hip/reduce.cuh", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/operators/gather_op.cuh", {"caffe2/operators/math/gather_op.cuh", "", CONV_INCLUDE, API_CAFFE2, UNSUPPORTED}}, + {"caffe2/core/common_cudnn.h", {"caffe2/core/hip/common_miopen.h", "", CONV_INCLUDE, API_CAFFE2}}, }; const std::map& CUDA_RENAMES_MAP() { @@ -88,5 +99,7 @@ const std::map& CUDA_RENAMES_MAP() { ret.insert(CUDA_FFT_FUNCTION_MAP.begin(), CUDA_FFT_FUNCTION_MAP.end()); ret.insert(CUDA_SPARSE_TYPE_NAME_MAP.begin(), CUDA_SPARSE_TYPE_NAME_MAP.end()); ret.insert(CUDA_SPARSE_FUNCTION_MAP.begin(), CUDA_SPARSE_FUNCTION_MAP.end()); + ret.insert(CUDA_CAFFE2_TYPE_NAME_MAP.begin(), CUDA_CAFFE2_TYPE_NAME_MAP.end()); + ret.insert(CUDA_CAFFE2_FUNCTION_MAP.begin(), CUDA_CAFFE2_FUNCTION_MAP.end()); return ret; }; diff --git a/projects/hip/hipify-clang/src/CUDA2HIP.h b/projects/hip/hipify-clang/src/CUDA2HIP.h index d95d4fea30..b8961097b3 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP.h +++ b/projects/hip/hipify-clang/src/CUDA2HIP.h @@ -61,6 +61,10 @@ extern const std::map CUDA_FFT_FUNCTION_MAP; extern const std::map CUDA_SPARSE_TYPE_NAME_MAP; // Maps the names of CUDA SPARSE API functions to the corresponding HIP functions extern const std::map CUDA_SPARSE_FUNCTION_MAP; +// Maps the names of CUDA CAFFE2 API types to the corresponding HIP types +extern const std::map CUDA_CAFFE2_TYPE_NAME_MAP; +// Maps the names of CUDA CAFFE2 API functions to the corresponding HIP functions +extern const std::map CUDA_CAFFE2_FUNCTION_MAP; /** * The union of all the above maps, except includes. diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_functions.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_functions.cpp new file mode 100644 index 0000000000..63860de262 --- /dev/null +++ b/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_functions.cpp @@ -0,0 +1,28 @@ +/* +Copyright (c) 2015 - present Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include "CUDA2HIP.h" + +// Maps the names of CUDA SPARSE API functions to the corresponding HIP functions +const std::map CUDA_CAFFE2_FUNCTION_MAP{ + {"cuda_stream", {"hip_stream", "", CONV_LIB_FUNC, API_CAFFE2}}, +}; \ No newline at end of file diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_types.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_types.cpp new file mode 100644 index 0000000000..4791cffeee --- /dev/null +++ b/projects/hip/hipify-clang/src/CUDA2HIP_CAFFE2_API_types.cpp @@ -0,0 +1,34 @@ +/* +Copyright (c) 2015 - present Advanced Micro Devices, Inc. All rights reserved. + +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: + +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. + +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include "CUDA2HIP.h" + +// Map of all types +const std::map CUDA_CAFFE2_TYPE_NAME_MAP{ + + // 5. Defines + {"REGISTER_CUDA_OPERATOR", {"REGISTER_HIP_OPERATOR", "", CONV_DEFINE, API_CAFFE2}}, + {"REGISTER_CUDA_OPERATOR_CREATOR", {"REGISTER_HIP_OPERATOR_CREATOR", "", CONV_DEFINE, API_CAFFE2}}, + + // 6. Classes + {"CUDAContext", {"HIPContext", "", CONV_TYPE, API_CAFFE2}}, +}; diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_Complex_API_functions.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_Complex_API_functions.cpp index af012e27cb..6e0c1a54e7 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_Complex_API_functions.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_Complex_API_functions.cpp @@ -22,7 +22,7 @@ THE SOFTWARE. #include "CUDA2HIP.h" -// Maps the names of CUDA Complex API types to the corresponding HIP types +// Maps the names of CUDA Complex API functions to the corresponding HIP functions const std::map CUDA_COMPLEX_FUNCTION_MAP{ {"cuCrealf", {"hipCrealf", "", CONV_COMPLEX, API_COMPLEX}}, {"cuCimagf", {"hipCimagf", "", CONV_COMPLEX, API_COMPLEX}}, diff --git a/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_functions.cpp b/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_functions.cpp index d89e292852..210ab1ee06 100644 --- a/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_functions.cpp +++ b/projects/hip/hipify-clang/src/CUDA2HIP_SPARSE_API_functions.cpp @@ -22,7 +22,7 @@ THE SOFTWARE. #include "CUDA2HIP.h" -// Maps the names of CUDA SPARSE API types to the corresponding HIP types +// Maps the names of CUDA SPARSE API functions to the corresponding HIP functions const std::map CUDA_SPARSE_FUNCTION_MAP{ // 5. cuSPARSE Helper Function Reference {"cusparseCreate", {"hipsparseCreate", "", CONV_LIB_FUNC, API_SPARSE}}, diff --git a/projects/hip/hipify-clang/src/Statistics.cpp b/projects/hip/hipify-clang/src/Statistics.cpp index 39f70e9d8a..e673957458 100644 --- a/projects/hip/hipify-clang/src/Statistics.cpp +++ b/projects/hip/hipify-clang/src/Statistics.cpp @@ -73,7 +73,8 @@ const char *apiNames[NUM_API_TYPES] = { "cuRAND API", "cuDNN API", "cuFFT API", - "cuSPARSE API" + "cuSPARSE API", + "CAFFE2 API" }; namespace { diff --git a/projects/hip/hipify-clang/src/Statistics.h b/projects/hip/hipify-clang/src/Statistics.h index 6d8986bc52..8305f0de93 100644 --- a/projects/hip/hipify-clang/src/Statistics.h +++ b/projects/hip/hipify-clang/src/Statistics.h @@ -131,6 +131,7 @@ enum ApiTypes { API_DNN, API_FFT, API_SPARSE, + API_CAFFE2, API_LAST }; constexpr int NUM_API_TYPES = (int) ApiTypes::API_LAST; diff --git a/projects/hip/tests/hipify-clang/lit.cfg b/projects/hip/tests/hipify-clang/lit.cfg index 0fd0aef275..00a3eeb873 100644 --- a/projects/hip/tests/hipify-clang/lit.cfg +++ b/projects/hip/tests/hipify-clang/lit.cfg @@ -14,6 +14,8 @@ lit_config.load_config(config, site_cfg) print("CUDA " + config.cuda_version + " will be used for testing.") config.excludes = ['cmdparser.hpp'] +config.excludes.append('spatial_batch_norm_op.h') +config.excludes.append('common_cudnn.h') if config.cuda_version_major == 7 and config.cuda_version_minor == 0: config.excludes.append('headers_test_09.cu') diff --git a/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/core/common_cudnn.h b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/core/common_cudnn.h new file mode 100644 index 0000000000..e9437c11f5 --- /dev/null +++ b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/core/common_cudnn.h @@ -0,0 +1,7 @@ +#ifndef CAFFE2_CORE_COMMON_CUDNN_H_ +#define CAFFE2_CORE_COMMON_CUDNN_H_ + +#include +#include + +#endif // CAFFE2_CORE_COMMON_CUDNN_H_ diff --git a/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/operators/spatial_batch_norm_op.h b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/operators/spatial_batch_norm_op.h new file mode 100644 index 0000000000..7b8a13788a --- /dev/null +++ b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2/operators/spatial_batch_norm_op.h @@ -0,0 +1,14 @@ +#ifndef CAFFE2_OPERATORS_SPATIAL_BATCH_NORM_OP_H_ +#define CAFFE2_OPERATORS_SPATIAL_BATCH_NORM_OP_H_ + +#include +#include +#include +#include +#include + +namespace caffe2 { + +} // namespace caffe2 + +#endif // CAFFE2_OPERATORS_SPATIAL_BATCH_NORM_OP_H_ diff --git a/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_01.cu b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_01.cu new file mode 100644 index 0000000000..3c82045d30 --- /dev/null +++ b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_01.cu @@ -0,0 +1,12 @@ +// RUN: %run_test hipify "%s" "%t" %hipify_args "-roc" %clang_args + +// NOTE: Nonworking code just for conversion testing + +// CHECK: #include +#include +#include +#include +// CHECK: #include "caffe2/operators/hip/spatial_batch_norm_op_miopen.hip" +#include "caffe2/operators/spatial_batch_norm_op.h" +// CHECK: #include "caffe2/core/hip/common_miopen.h" +#include "caffe2/core/common_cudnn.h" diff --git a/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_02.cu b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_02.cu new file mode 100644 index 0000000000..bcbe440a15 --- /dev/null +++ b/projects/hip/tests/hipify-clang/unit_tests/libraries/CAFFE2/caffe2_02.cu @@ -0,0 +1,103 @@ +// RUN: %run_test hipify "%s" "%t" %hipify_args "-roc" %clang_args + +// NOTE: Nonworking code just for conversion testing + +// CHECK: #include +#include +#include +#include +#include + +namespace caffe2 { + +// Operator Definition. +struct OperatorDef { + int input = 1; + int output = 2; + int name = 3; +}; + +class OperatorBase; +class Workspace; + +template +class Observable { + public: + Observable() = default; + + Observable(Observable&&) = default; + Observable& operator =(Observable&&) = default; + + virtual ~Observable() = default; +}; + +template +class ObserverBase { + public: + explicit ObserverBase(T* subject) : subject_(subject) {} + + virtual void Start() {} + virtual void Stop() {} + + virtual std::string debugInfo() { + return "Not implemented."; + } + + virtual ~ObserverBase() noexcept {}; + + T* subject() const { + return subject_; + } + + protected: + T* subject_; +}; + +typedef ObserverBase OperatorObserver; + +class OperatorBase : public Observable { + public: + explicit OperatorBase(const OperatorDef& operator_def, Workspace* ws); + virtual ~OperatorBase() noexcept {} +}; + +template +class Operator : public OperatorBase { + public: + explicit Operator(const OperatorDef& operator_def, Workspace* ws) + : OperatorBase(operator_def, ws), context_(operator_def.device_option()) { + context_.SwitchToDevice(); + } + ~Operator() noexcept override {} +}; + +template +class DummyEmptyOp : public Operator { + public: + DummyEmptyOp(const OperatorDef& def, Workspace* ws) + : Operator(def, ws) {} + + bool RunOnDevice() final { return true; } +}; + + +class CUDAContext { +public: + CUDAContext(); + virtual ~CUDAContext() noexcept {} +}; + +#define REGISTER_CUDA_OPERATOR(name, ...) \ + void CAFFE2_PLEASE_ADD_OPERATOR_SCHEMA_FOR_##name(); \ + static void CAFFE_ANONYMOUS_VARIABLE_CUDA##name() { \ + CAFFE2_PLEASE_ADD_OPERATOR_SCHEMA_FOR_##name(); \ + } + +#define REGISTER_CUDA_OPERATOR_CREATOR(key, ...) + +// CHECK: REGISTER_HIP_OPERATOR(Operator, DummyEmptyOp); +REGISTER_CUDA_OPERATOR(Operator, DummyEmptyOp); +// CHECK: REGISTER_HIP_OPERATOR_CREATOR(Operator, DummyEmptyOp); +REGISTER_CUDA_OPERATOR_CREATOR(Operator, DummyEmptyOp); + +}