Merge remote-tracking branch 'origin/hiprng-mlopen' into hip

Conflicts:
	include/caffe/util/device_alternate.hpp

Author: whchung

Makefile

@@ -176,7 +176,7 @@ HIP_LIB_DIR += /usr/local/cuda/lib64
 
 ifneq (, $(findstring hcc, $(HIP_PLATFORM)))
 	#HIP_LIBS := hip_hcc hcblas
-	HIP_LIBS := hcblas
+	HIP_LIBS := hcblas hcrng
 else ifneq (, $(findstring nvcc, $(HIP_PLATFORM)))
 	HIP_LIBS := cudart cublas curand
 endif

Makefile.config

@@ -104,8 +104,8 @@ PYTHON_LIB := /usr/lib
 # WITH_PYTHON_LAYER := 1
 
 # Whatever else you find you need goes here.
-INCLUDE_DIRS := $(PYTHON_INCLUDE) /usr/local/include /opt/rocm/hip/include /opt/rocm/hcblas/include
-LIBRARY_DIRS := $(PYTHON_LIB) /usr/local/lib /usr/lib /opt/rocm/hcblas/lib
+INCLUDE_DIRS := $(PYTHON_INCLUDE) /usr/local/include /opt/rocm/hip/include /opt/rocm/profiler/CXLActivityLogger/include /opt/rocm/hcblas/include /opt/rocm/hcrng/include
+LIBRARY_DIRS := $(PYTHON_LIB) /usr/local/lib /usr/lib /opt/rocm/hcblas/lib /opt/rocm/hcrng/lib
 
 # If Homebrew is installed at a non standard location (for example your home directory) and you use it for general dependencies
 # INCLUDE_DIRS += $(shell brew --prefix)/include

README.amd-tips

@@ -1,3 +1,8 @@
+# Build caffe:
+$ source setup.sh
+$ make DEBUG=1 -j20
+$ build/tools/caffe train --solver=examples/cifar10/cifar10_quick_solver.prototxt  --gpu 0,1
+
 
 ### CAFFE directed tests: 
 # Run all tests:

README.md

@@ -115,7 +115,7 @@ D. Unit Testing
 
 a. Support libraries 
 
-          sudo apt-get install libprotobuf-dev libleveldb-dev libsnappy-dev libopencv-dev libhdf5-serial-dev protobuf-compiler libatlas-base-dev libblas-dev libgflags-dev libgoogle-glog-dev liblmdb-dev libboost-all-dev
+          sudo apt-get install libprotobuf-dev libleveldb-dev libsnappy-dev libopencv-dev libhdf5-serial-dev protobuf-compiler libatlas-base-dev libblas-dev libgflags-dev libgoogle-glog-dev liblmdb-dev libboost-all-dev python-h5py
 
 b. HcBLAS library
 
@@ -138,4 +138,4 @@ To improve build time, one could as well invoke make -j <number of threads>
 
 After done with A, B and C, Now its time to test. Run the following commands to perform unit testing of different components of Caffe.
 
-             ./build/test/test_all.testbin
+             ./build/test/test_all.testbin

include/caffe/common.hpp

@@ -98,7 +98,7 @@ using std::vector;
 void GlobalInit(int* pargc, char*** pargv);
 
 // A singleton class to hold common caffe stuff, such as the handler that
-// caffe is going to use for hipblas, curand, etc.
+// caffe is going to use for hipblas, hiprng, etc.
 class Caffe {
  public:
   ~Caffe();
@@ -124,7 +124,7 @@ class Caffe {
     shared_ptr<Generator> generator_;
   };
 
-  // Getters for boost rng, curand, and hipblas handles
+  // Getters for boost rng, hiprng, and hipblas handles
   inline static RNG& rng_stream() {
     if (!Get().random_generator_) {
       Get().random_generator_.reset(new RNG());
@@ -133,10 +133,9 @@ class Caffe {
   }
 #ifndef CPU_ONLY
   inline static hipblasHandle_t hipblas_handle() { return Get().hipblas_handle_; }
-  // TODO HIP Equivalent
-  /*inline static curandGenerator_t curand_generator() {
-    return Get().curand_generator_;
-  }*/
+  inline static hiprngGenerator_t hiprng_generator() {
+    return Get().hiprng_generator_;
+  }
 #endif
 
   // Returns the mode: running on CPU or GPU.
@@ -147,9 +146,9 @@ class Caffe {
   // freed in a non-pinned way, which may cause problems - I haven't verified
   // it personally but better to note it here in the header file.
   inline static void set_mode(Brew mode) { Get().mode_ = mode; }
-  // Sets the random seed of both boost and curand
+  // Sets the random seed of both boost and hiprng
   static void set_random_seed(const unsigned int seed);
-  // Sets the device. Since we have hipblas and curand stuff, set device also
+  // Sets the device. Since we have hipblas and hiprng stuff, set device also
   // requires us to reset those values.
   static void SetDevice(const int device_id);
   // Prints the current GPU status.
@@ -168,8 +167,7 @@ class Caffe {
  protected:
 #ifndef CPU_ONLY
   hipblasHandle_t hipblas_handle_;
-  //TODO: HIP Equivalent
-  //curandGenerator_t curand_generator_;
+  hiprngGenerator_t hiprng_generator_;
 #endif
   shared_ptr<RNG> random_generator_;
 

include/caffe/util/device_alternate.hpp

@@ -43,6 +43,7 @@ void classname<Dtype>::funcname##_##gpu(const vector<Blob<Dtype>*>& top, \
 #endif
 
 #include <hipblas.h>
+#include <hiprng.h>
 #ifdef USE_ACCMI
 #include "caffe/util/cudnn.hpp"
 #endif
@@ -69,14 +70,12 @@ void classname<Dtype>::funcname##_##gpu(const vector<Blob<Dtype>*>& top, \
       << caffe::hipblasGetErrorString(status); \
   } while (0)
 
-// TODO: Get HIP equivalent
-/*#define CURAND_CHECK(condition) \
+#define HIPRNG_CHECK(condition) \
   do { \
-    curandStatus_t status = condition; \
-    CHECK_EQ(status, CURAND_STATUS_SUCCESS) << " " \
-      << caffe::curandGetErrorString(status); \
+    hiprngStatus_t status = condition; \
+    CHECK_EQ(status,HIPRNG_STATUS_SUCCESS) << " " \
+      << caffe::hiprngGetErrorString(status); \
   } while (0)
-*/
 
 // HIP: grid stride looping
 #define HIP_KERNEL_LOOP(i, n) \
@@ -85,13 +84,13 @@ void classname<Dtype>::funcname##_##gpu(const vector<Blob<Dtype>*>& top, \
        i += hipBlockDim_x * hipGridDim_x)
 
 // HIP: check for error after kernel execution and exit loudly if there is one.
-//TODO: Get HIP equivalent
-//#define HIP_POST_KERNEL_CHECK HIP_CHECK(cudaPeekAtLastError())
+#define HIP_POST_KERNEL_CHECK HIP_CHECK(hipPeekAtLastError())
 
 namespace caffe {
 
 // HIP: library error reporting.
 const char* hipblasGetErrorString(hipblasStatus_t error);
+const char* hiprngGetErrorString(hiprngStatus_t error);
 // HIP: use 512 threads per block
 
 #ifdef __HIP_PLATFORM_NVCC__

include/caffe/util/math_functions.hpp

@@ -217,9 +217,9 @@ void caffe_gpu_rng_uniform(const int n, unsigned int* r);
 
 // caffe_gpu_rng_uniform with four arguments generates floats in the range
 // (a, b] (strictly greater than a, less than or equal to b) due to the
-// specification of curandGenerateUniform.  With a = 0, b = 1, just calls
-// curandGenerateUniform; with other limits will shift and scale the outputs
-// appropriately after calling curandGenerateUniform.
+// specification of hiprngGenerateUniform.  With a = 0, b = 1, just calls
+// hiprngGenerateUniform; with other limits will shift and scale the outputs
+// appropriately after calling hiprngGenerateUniform.
 template <typename Dtype>
 void caffe_gpu_rng_uniform(const int n, const Dtype a, const Dtype b, Dtype* r);
 

src/caffe/common.cpp

@@ -106,47 +106,46 @@ void* Caffe::RNG::generator() {
 
 Caffe::Caffe() 
   // TODO: HIP Equivalent
-   : hipblas_handle_(NULL), random_generator_(),
+   : hipblas_handle_(NULL), hiprng_generator_(NULL), random_generator_(),
     mode_(Caffe::CPU), solver_count_(1), root_solver_(true) {
   // Try to create a hipblas handler, and report an error if failed (but we will
   // keep the program running as one might just want to run CPU code).
   if (hipblasCreate(&hipblas_handle_) != HIPBLAS_STATUS_SUCCESS) {
     LOG(ERROR) << "Cannot create Cublas handle. Cublas won't be available.";
   }
-  // Try to create a curand handler.
-  /*if (curandCreateGenerator(&curand_generator_, CURAND_RNG_PSEUDO_DEFAULT)
-      != CURAND_STATUS_SUCCESS ||
-      curandSetPseudoRandomGeneratorSeed(curand_generator_, cluster_seedgen())
-      != CURAND_STATUS_SUCCESS) {
+  // Try to create a hiprng handler.
+  if (hiprngCreateGenerator(&hiprng_generator_, HIPRNG_RNG_PSEUDO_MRG32K3A)
+      != HIPRNG_STATUS_SUCCESS ||
+      hiprngSetPseudoRandomGeneratorSeed(hiprng_generator_, cluster_seedgen())
+      != HIPRNG_STATUS_SUCCESS) {
     LOG(ERROR) << "Cannot create Curand generator. Curand won't be available.";
-  }*/
+  }
 }
 
 Caffe::~Caffe() {
-  // TODO: HIP Equivalent
   if (hipblas_handle_) HIPBLAS_CHECK(hipblasDestroy(hipblas_handle_));
-  /*if (curand_generator_) {
-    CURAND_CHECK(curandDestroyGenerator(curand_generator_));
-  }*/
+  if (hiprng_generator_) {
+    HIPRNG_CHECK(hiprngDestroyGenerator(hiprng_generator_));
+  }
 }
 
 void Caffe::set_random_seed(const unsigned int seed) {
   // Curand seed
-  // TODO HIP Equivalent
-  /*static bool g_curand_availability_logged = false;
-  if (Get().curand_generator_) {
-    CURAND_CHECK(curandSetPseudoRandomGeneratorSeed(curand_generator(),
-        seed));
-    CURAND_CHECK(curandSetGeneratorOffset(curand_generator(), 0));
+  static bool g_hiprng_availability_logged = false;
+  if (Get().hiprng_generator_) {
+    HIPRNG_CHECK(hiprngSetPseudoRandomGeneratorSeed(hiprng_generator(),
+      seed));
+    // TODO: support in HIP equivalent
+    //HIPRNG_CHECK(hiprngSetGeneratorOffset(hiprng_generator(), 0));
   } else {
-    if (!g_curand_availability_logged) {
+    if (!g_hiprng_availability_logged) {
         LOG(ERROR) <<
-            "Curand not available. Skipping setting the curand seed.";
-        g_curand_availability_logged = true;
+            "Curand not available. Skipping setting the hiprng seed.";
+        g_hiprng_availability_logged = true;
     }
   }
   // RNG seed
-  Get().random_generator_.reset(new RNG(seed));*/
+  Get().random_generator_.reset(new RNG(seed));
 }
 
 void Caffe::SetDevice(const int device_id) {
@@ -158,16 +157,15 @@ void Caffe::SetDevice(const int device_id) {
   // The call to hipSetDevice must come before any calls to Get, which
   // may perform initialization using the GPU.
   HIP_CHECK(hipSetDevice(device_id));
-  //TODO HIP equivalent
   if (Get().hipblas_handle_) HIPBLAS_CHECK(hipblasDestroy(Get().hipblas_handle_));
-  /*if (Get().curand_generator_) {
-    CURAND_CHECK(curandDestroyGenerator(Get().curand_generator_));
-  }*/
+  if (Get().hiprng_generator_) {
+    HIPRNG_CHECK(hiprngDestroyGenerator(Get().hiprng_generator_));
+  }
   HIPBLAS_CHECK(hipblasCreate(&Get().hipblas_handle_));
-  /*CURAND_CHECK(curandCreateGenerator(&Get().curand_generator_,
-      CURAND_RNG_PSEUDO_DEFAULT));
-  CURAND_CHECK(curandSetPseudoRandomGeneratorSeed(Get().curand_generator_,
-      cluster_seedgen()));*/
+  HIPRNG_CHECK(hiprngCreateGenerator(&Get().hiprng_generator_,
+      HIPRNG_RNG_PSEUDO_MRG32K3A));
+  HIPRNG_CHECK(hiprngSetPseudoRandomGeneratorSeed(Get().hiprng_generator_,
+      cluster_seedgen()));
 }
 
 void Caffe::DeviceQuery() {
@@ -263,7 +261,7 @@ void* Caffe::RNG::generator() {
 
 
 const char* hipblasGetErrorString(hipblasStatus_t error) {
-  /*switch (error) {
+  switch (error) {
   case HIPBLAS_STATUS_SUCCESS:
     return "HIPBLAS_STATUS_SUCCESS";
   case HIPBLAS_STATUS_NOT_INITIALIZED:
@@ -278,46 +276,49 @@ const char* hipblasGetErrorString(hipblasStatus_t error) {
     return "HIPBLAS_STATUS_EXECUTION_FAILED";
   case HIPBLAS_STATUS_INTERNAL_ERROR:
     return "HIPBLAS_STATUS_INTERNAL_ERROR";
+  case HIPBLAS_STATUS_NOT_SUPPORTED:
+    return "HIPBLAS_STATUS_NOT_SUPPORTED";
 #if HIP_VERSION >= 6000
   case HIPBLAS_STATUS_INTERNAL_ERROR:
     return "HIPBLAS_STATUS_INTERNAL_ERROR";
 #endif
-  }*/
+  }
   return "Unknown hipblas status";
 }
 
-// TODO HIP Equivalent
-/*const char* curandGetErrorString(curandStatus_t error) {
+const char* hiprngGetErrorString(hiprngStatus_t error) {
   switch (error) {
-  case CURAND_STATUS_SUCCESS:
-    return "CURAND_STATUS_SUCCESS";
-  case CURAND_STATUS_VERSION_MISMATCH:
-    return "CURAND_STATUS_VERSION_MISMATCH";
-  case CURAND_STATUS_NOT_INITIALIZED:
-    return "CURAND_STATUS_NOT_INITIALIZED";
-  case CURAND_STATUS_ALLOCATION_FAILED:
-    return "CURAND_STATUS_ALLOCATION_FAILED";
-  case CURAND_STATUS_TYPE_ERROR:
-    return "CURAND_STATUS_TYPE_ERROR";
-  case CURAND_STATUS_OUT_OF_RANGE:
-    return "CURAND_STATUS_OUT_OF_RANGE";
-  case CURAND_STATUS_LENGTH_NOT_MULTIPLE:
-    return "CURAND_STATUS_LENGTH_NOT_MULTIPLE";
-  case CURAND_STATUS_DOUBLE_PRECISION_REQUIRED:
-    return "CURAND_STATUS_DOUBLE_PRECISION_REQUIRED";
-  case CURAND_STATUS_LAUNCH_FAILURE:
-    return "CURAND_STATUS_LAUNCH_FAILURE";
-  case CURAND_STATUS_PREEXISTING_FAILURE:
-    return "CURAND_STATUS_PREEXISTING_FAILURE";
-  case CURAND_STATUS_INITIALIZATION_FAILED:
-    return "CURAND_STATUS_INITIALIZATION_FAILED";
-  case CURAND_STATUS_ARCH_MISMATCH:
-    return "CURAND_STATUS_ARCH_MISMATCH";
-  case CURAND_STATUS_INTERNAL_ERROR:
-    return "CURAND_STATUS_INTERNAL_ERROR";
+  case HIPRNG_STATUS_INVALID_STREAM_CREATOR:
+    return "HIPRNG_STATUS_INVALID_STREAM_CREATOR";
+  case HIPRNG_STATUS_SUCCESS:
+    return "HIPRNG_STATUS_SUCCESS";
+  case HIPRNG_STATUS_VERSION_MISMATCH:
+    return "HIPRNG_STATUS_VERSION_MISMATCH";
+  //case HIPRNG_STATUS_NOT_INITIALIZED:
+    //return "HIPRNG_STATUS_NOT_INITIALIZED";
+  case HIPRNG_STATUS_ALLOCATION_FAILED:
+    return "HIPRNG_STATUS_ALLOCATION_FAILED";
+  case HIPRNG_STATUS_TYPE_ERROR:
+    return "HIPRNG_STATUS_TYPE_ERROR";
+  //case HIPRNG_STATUS_OUT_OF_RANGE:
+    //return "HIPRNG_STATUS_OUT_OF_RANGE";
+  //case HIPRNG_STATUS_LENGTH_NOT_MULTIPLE:
+    //return "HIPRNG_STATUS_LENGTH_NOT_MULTIPLE";
+  //case HIPRNG_STATUS_DOUBLE_PRECISION_REQUIRED:
+    //return "HIPRNG_STATUS_DOUBLE_PRECISION_REQUIRED";
+  //case HIPRNG_STATUS_LAUNCH_FAILURE:
+    //return "HIPRNG_STATUS_LAUNCH_FAILURE";
+  //case HIPRNG_STATUS_PREEXISTING_FAILURE:
+    //return "HIPRNG_STATUS_PREEXISTING_FAILURE";
+  case HIPRNG_STATUS_INITIALIZATION_FAILED:
+    return "HIPRNG_STATUS_INITIALIZATION_FAILED";
+  //case HIPRNG_STATUS_ARCH_MISMATCH:
+    //return "HIPRNG_STATUS_ARCH_MISMATCH";
+  //case HIPRNG_STATUS_INTERNAL_ERROR:
+    //return "HIPRNG_STATUS_INTERNAL_ERROR";
   }
-  return "Unknown curand status";
-}*/
+  return "Unknown hiprng status";
+}
 
 #endif  // CPU_ONLY
 

src/caffe/test/CMakeLists.txt

@@ -10,10 +10,10 @@ caffe_leave_only_selected_tests(test_cuda ${BUILD_only_tests})
 # That's why the lines below are commented. TODO: remove them
 
 # definition needed to include CMake generated files
-#add_definitions(-DCMAKE_BUILD)
+add_definitions(-DCMAKE_BUILD)
 
 # generates test_data/sample_data_list.txt.gen.cmake
-#caffe_configure_testdatafile(test_data/sample_data_list.txt)
+caffe_configure_testdatafile(test_data/sample_data_list.txt)
 
 set(the_target test.testbin)
 set(test_args --gtest_shuffle)

src/caffe/test/test_common.cpp

@@ -15,8 +15,7 @@ class CommonTest : public ::testing::Test {};
 TEST_F(CommonTest, TestCublasHandlerGPU) {
   int hip_device_id;
   HIP_CHECK(hipGetDevice(&hip_device_id));
-  //TODO: HIP equivalent
-  //EXPECT_TRUE(Caffe::cublas_handle());
+  EXPECT_TRUE(Caffe::hipblas_handle());
 }
 
 #endif
@@ -49,16 +48,15 @@ TEST_F(CommonTest, TestRandSeedGPU) {
   SyncedMemory data_a(10 * sizeof(unsigned int));
   SyncedMemory data_b(10 * sizeof(unsigned int));
   Caffe::set_random_seed(1701);
-  // TODO: HIP Equivalent
-  /*CURAND_CHECK(curandGenerate(Caffe::curand_generator(),
+  HIPRNG_CHECK(hiprngGenerate(Caffe::hiprng_generator(),
         static_cast<unsigned int*>(data_a.mutable_gpu_data()), 10));
   Caffe::set_random_seed(1701);
-  CURAND_CHECK(curandGenerate(Caffe::curand_generator(),
+  HIPRNG_CHECK(hiprngGenerate(Caffe::hiprng_generator(),
         static_cast<unsigned int*>(data_b.mutable_gpu_data()), 10));
   for (int i = 0; i < 10; ++i) {
     EXPECT_EQ(((const unsigned int*)(data_a.cpu_data()))[i],
         ((const unsigned int*)(data_b.cpu_data()))[i]);
-  }*/
+  }
 }
 
 #endif