[HIPIFY][Caffe2] Initial Caffe2 support

[ROCm/hip commit: 2619f22e5c]

projects/hip/hipify-clang/src/CUDA2HIP.cpp

@@ -64,6 +64,17 @@ const std::map <llvm::StringRef, hipCounter> 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<llvm::StringRef, hipCounter>& CUDA_RENAMES_MAP() {
@@ -88,5 +99,7 @@ const std::map<llvm::StringRef, hipCounter>& 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;
 };

projects/hip/hipify-clang/src/CUDA2HIP.h

@@ -61,6 +61,10 @@ extern const std::map<llvm::StringRef, hipCounter> CUDA_FFT_FUNCTION_MAP;
 extern const std::map<llvm::StringRef, hipCounter> CUDA_SPARSE_TYPE_NAME_MAP;
 // Maps the names of CUDA SPARSE API functions to the corresponding HIP functions
 extern const std::map<llvm::StringRef, hipCounter> CUDA_SPARSE_FUNCTION_MAP;
+// Maps the names of CUDA CAFFE2 API types to the corresponding HIP types
+extern const std::map<llvm::StringRef, hipCounter> CUDA_CAFFE2_TYPE_NAME_MAP;
+// Maps the names of CUDA CAFFE2 API functions to the corresponding HIP functions
+extern const std::map<llvm::StringRef, hipCounter> CUDA_CAFFE2_FUNCTION_MAP;
 
 /**
   * The union of all the above maps, except includes.

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<llvm::StringRef, hipCounter> CUDA_CAFFE2_FUNCTION_MAP{
+  {"cuda_stream",                                   {"hip_stream",                               "", CONV_LIB_FUNC, API_CAFFE2}},
+};

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<llvm::StringRef, hipCounter> 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}},
+};

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<llvm::StringRef, hipCounter> CUDA_COMPLEX_FUNCTION_MAP{
   {"cuCrealf",               {"hipCrealf",               "", CONV_COMPLEX, API_COMPLEX}},
   {"cuCimagf",               {"hipCimagf",               "", CONV_COMPLEX, API_COMPLEX}},

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<llvm::StringRef, hipCounter> CUDA_SPARSE_FUNCTION_MAP{
   // 5. cuSPARSE Helper Function Reference
   {"cusparseCreate",                              {"hipsparseCreate",                              "", CONV_LIB_FUNC, API_SPARSE}},

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 {

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;

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')

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 <array>
+#include <mutex>
+
+#endif  // CAFFE2_CORE_COMMON_CUDNN_H_

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 <algorithm>
+#include <array>
+#include <functional>
+#include <string>
+#include <vector>
+
+namespace caffe2 {
+
+} // namespace caffe2
+
+#endif // CAFFE2_OPERATORS_SPATIAL_BATCH_NORM_OP_H_

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 <hip/hip_runtime.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <math.h>
+// 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"

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 <hip/hip_runtime.h>
+#include <stdio.h>
+#include <stdlib.h>
+#include <math.h>
+#include <string>
+
+namespace caffe2 {
+
+// Operator Definition.
+struct OperatorDef {
+  int input = 1;
+  int output = 2;
+  int name = 3;
+};
+
+class OperatorBase;
+class Workspace;
+
+template <class T>
+class Observable {
+ public:
+  Observable() = default;
+
+  Observable(Observable&&) = default;
+  Observable& operator =(Observable&&) = default;
+
+  virtual ~Observable() = default;
+};
+
+template <class T>
+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<OperatorBase> OperatorObserver;
+
+class OperatorBase : public Observable<OperatorBase> {
+ public:
+  explicit OperatorBase(const OperatorDef& operator_def, Workspace* ws);
+  virtual ~OperatorBase() noexcept {}
+};
+
+template <class Context>
+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 Context>
+class DummyEmptyOp : public Operator<Context> {
+ public:
+  DummyEmptyOp(const OperatorDef& def, Workspace* ws)
+      : Operator<Context>(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<HIPContext>);
+REGISTER_CUDA_OPERATOR(Operator, DummyEmptyOp<CUDAContext>);
+// CHECK: REGISTER_HIP_OPERATOR_CREATOR(Operator, DummyEmptyOp<HIPContext>);
+REGISTER_CUDA_OPERATOR_CREATOR(Operator, DummyEmptyOp<CUDAContext>);
+
+}