Merge pull request #279 from sony/feature/20210107-add_seed_function

Add seed function

include/nbla/cuda/cuda.hpp

@@ -56,7 +56,7 @@ class NBLA_CUDA_API Cuda : public BackendBase {
   std::shared_ptr<cudaEvent_t> cuda_event(unsigned int flags, int device = -1);
 
   /** Get cuRAND global generator **/
-  curandGenerator_t curand_generator();
+  curandGenerator_t &curand_generator();
 
   void curand_set_seed(int seed);
 
@@ -136,7 +136,8 @@ class NBLA_CUDA_API Cuda : public BackendBase {
   unordered_map<int, curandGenerator_t> curand_generators_;
   unordered_map<int, unordered_map<unsigned int, vector<cudaEvent_t>>>
       cuda_unused_events_;
-  vector<string> array_classes_; ///< Available array classes
+  vector<string> array_classes_;        ///< Available array classes
+  unordered_map<int, int> seed_counts_; /// this is used to check seed update
 
   /*
     NOTE: Allocators must be retained as shared_ptr in order to be passed to a

include/nbla/cuda/function/dropout.hpp

@@ -39,8 +39,6 @@ template <typename T> class DropoutCuda : public Dropout<T> {
     if (this->seed_ != -1) {
       // CURAND_RNG_PSEUDO_DEFAULT is CURAND_RNG_PSEUDO_XORWOW.
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~DropoutCuda() {

include/nbla/cuda/function/rand.hpp

@@ -37,8 +37,6 @@ template <typename T> class RandCuda : public Rand<T> {
     cuda_set_device(device_);
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandCuda() {

include/nbla/cuda/function/randint.hpp

@@ -37,8 +37,6 @@ template <typename T> class RandintCuda : public Randint<T> {
     cuda_set_device(device_);
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandintCuda() {

include/nbla/cuda/function/randn.hpp

@@ -36,8 +36,6 @@ template <typename T> class RandnCuda : public Randn<T> {
         device_(std::stoi(ctx.device_id)) {
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandnCuda() {

include/nbla/cuda/function/random_choice.hpp

@@ -32,8 +32,6 @@ template <typename T> class RandomChoiceCuda : public RandomChoice<T> {
     cuda_set_device(device_);
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandomChoiceCuda() {

include/nbla/cuda/function/random_crop.hpp

@@ -33,8 +33,6 @@ template <typename T> class RandomCropCuda : public RandomCrop<T> {
     cuda_set_device(std::stoi(ctx.device_id));
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandomCropCuda() {}

include/nbla/cuda/function/random_erase.hpp

@@ -39,8 +39,6 @@ template <typename T> class RandomEraseCuda : public RandomErase<T> {
     cuda_set_device(device_);
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandomEraseCuda() {

include/nbla/cuda/function/random_flip.hpp

@@ -34,8 +34,6 @@ template <typename T> class RandomFlipCuda : public RandomFlip<T> {
     cuda_set_device(std::stoi(ctx.device_id));
     if (this->seed_ != -1) {
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
   virtual ~RandomFlipCuda() {}

src/nbla/cuda/cuda.cpp

@@ -22,6 +22,7 @@
 #include <nbla/memory/caching_allocator_with_buckets.hpp>
 #include <nbla/memory/naive_allocator.hpp>
 #include <nbla/memory/virtual_caching_allocator.hpp>
+#include <nbla/random_manager.hpp>
 
 namespace nbla {
 
@@ -220,17 +221,29 @@ shared_ptr<cudaStream_t> Cuda::get_stream(unsigned int flags,
   }
 }
 
-curandGenerator_t Cuda::curand_generator() {
+curandGenerator_t &Cuda::curand_generator() {
   // Get current device
   int device = cuda_get_device();
   std::lock_guard<decltype(mtx_curand_)> lock(mtx_curand_);
   // Find device rng
   auto it = this->curand_generators_.find(device);
+  // Get the latest RandomManager states
+  int seed_count = SingletonManager::get<RandomManager>()->get_count();
+  int seed = SingletonManager::get<RandomManager>()->get_seed();
   // Create a new one
   if (it == this->curand_generators_.end()) {
-    curandGenerator_t gen = curand_create_generator();
+    curandGenerator_t gen = curand_create_generator(seed);
     this->curand_generators_.insert({device, gen});
-    return gen;
+    this->seed_counts_.insert({device, seed_count});
+    return this->curand_generators_[device];
+  } else if (this->seed_counts_[it->first] < seed_count) {
+    // Destroy old generator
+    curand_destroy_generator(it->second);
+    // Recreate
+    curandGenerator_t gen = curand_create_generator(seed);
+    this->curand_generators_[device] = gen;
+    this->seed_counts_[device] = seed_count;
+    return this->curand_generators_[device];
   }
   return it->second;
 }

src/nbla/cuda/cudnn/function/generic/gru.cu

@@ -17,6 +17,7 @@
 #include <nbla/cuda/common.hpp>
 #include <nbla/cuda/cudnn/cudnn.hpp>
 #include <nbla/cuda/cudnn/function/gru.hpp>
+#include <nbla/random_manager.hpp>
 #include <nbla/variable.hpp>
 
 #include <array>
@@ -405,12 +406,12 @@ void GRUCudaCudnn<T>::setup_impl(const Variables &inputs,
   state_array_.reshape(Shape_t{static_cast<Size_t>(dropout_stateSize)}, true);
   void *state_ptr =
       state_array_.cast(dtypes::BYTE, this->ctx_, true)->pointer<void>();
-  std::random_device seed_gen;
-  std::default_random_engine engine(seed_gen());
+  std::mt19937 &rgen =
+      SingletonManager::get<RandomManager>()->get_rand_generator();
   std::uniform_int_distribution<> dist(0, 999);
   NBLA_CUDNN_CHECK(cudnnSetDropoutDescriptor(dropout_desc_.desc, cudnn_handle,
                                              this->dropout_, state_ptr,
-                                             dropout_stateSize, dist(engine)));
+                                             dropout_stateSize, dist(rgen)));
 
 // Set RNN descriptor.
 #if CUDNN_VERSION >= 7000

src/nbla/cuda/cudnn/function/generic/lstm.cu

@@ -17,6 +17,7 @@
 #include <nbla/cuda/common.hpp>
 #include <nbla/cuda/cudnn/cudnn.hpp>
 #include <nbla/cuda/cudnn/function/lstm.hpp>
+#include <nbla/random_manager.hpp>
 #include <nbla/variable.hpp>
 
 #include <array>
@@ -413,12 +414,12 @@ void LSTMCudaCudnn<T>::setup_impl(const Variables &inputs,
   state_array_.reshape(Shape_t{static_cast<Size_t>(dropout_stateSize)}, true);
   void *state_ptr =
       state_array_.cast(dtypes::BYTE, this->ctx_, true)->pointer<void>();
-  std::random_device seed_gen;
-  std::default_random_engine engine(seed_gen());
+  std::mt19937 &rgen =
+      SingletonManager::get<RandomManager>()->get_rand_generator();
   std::uniform_int_distribution<> dist(0, 999);
   NBLA_CUDNN_CHECK(cudnnSetDropoutDescriptor(dropout_desc_.desc, cudnn_handle,
                                              this->dropout_, state_ptr,
-                                             dropout_stateSize, dist(engine)));
+                                             dropout_stateSize, dist(rgen)));
   state_array_.array()->clear();
 
 // Set RNN descriptor.

src/nbla/cuda/cudnn/function/generic/rnn.cu

@@ -17,6 +17,7 @@
 #include <nbla/cuda/common.hpp>
 #include <nbla/cuda/cudnn/cudnn.hpp>
 #include <nbla/cuda/cudnn/function/rnn.hpp>
+#include <nbla/random_manager.hpp>
 #include <nbla/variable.hpp>
 
 #include <array>
@@ -397,12 +398,12 @@ void RNNCudaCudnn<T>::setup_impl(const Variables &inputs,
   state_array_.reshape(Shape_t{static_cast<Size_t>(dropout_stateSize)}, true);
   void *state_ptr =
       state_array_.cast(dtypes::BYTE, this->ctx_, true)->pointer<void>();
-  std::random_device seed_gen;
-  std::default_random_engine engine(seed_gen());
+  std::mt19937 &rgen =
+      SingletonManager::get<RandomManager>()->get_rand_generator();
   std::uniform_int_distribution<> dist(0, 999);
   NBLA_CUDNN_CHECK(cudnnSetDropoutDescriptor(dropout_desc_.desc, cudnn_handle,
                                              this->dropout_, state_ptr,
-                                             dropout_stateSize, dist(engine)));
+                                             dropout_stateSize, dist(rgen)));
 
 // Set RNN descriptor.
 #if CUDNN_VERSION >= 7000

src/nbla/cuda/function/generic/dropout.cu

@@ -54,8 +54,10 @@ void DropoutCuda<T>::forward_impl(const Variables &inputs,
   Tc *y = outputs[0]->cast_data_and_get_pointer<Tc>(this->ctx_, true);
   Variable &mask = this->mask_;
   float *m = mask.cast_data_and_get_pointer<float>(this->ctx_, true);
-  curand_generate_rand<float>(curand_generator_, 0.0f, 1.0f, m,
-                              inputs[0]->size());
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<float>(gen, 0.0f, 1.0f, m, inputs[0]->size());
   NBLA_CUDA_LAUNCH_KERNEL_SIMPLE(kernel_dropout_forward, inputs[0]->size(),
                                  this->scale_, this->p_, x, y, m);
 }

src/nbla/cuda/function/generic/inq_affine.cu

@@ -147,8 +147,6 @@ void INQAffineCuda<T, T1>::setup_impl(const Variables &inputs,
     if (this->seed_ != -1) {
       // CURAND_RNG_PSEUDO_DEFAULT is CURAND_RNG_PSEUDO_XORWOW.
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
 
@@ -249,7 +247,11 @@ void INQAffineCuda<T, T1>::forward_impl(const Variables &inputs,
       } else {
         // random selection (we re-use old_weights here to keep the random
         // values)
-        curand_generate_rand<Tc>(curand_generator_, 0.0f, 1.0f, old_weights,
+        curandGenerator_t &gen =
+            this->seed_ == -1
+                ? SingletonManager::get<Cuda>()->curand_generator()
+                : curand_generator_;
+        curand_generate_rand<Tc>(gen, 0.0f, 1.0f, old_weights,
                                  inputs[0]->size());
         NBLA_CUDA_LAUNCH_KERNEL_SIMPLE((kernel_random_selection<Tc, T1>),
                                        inputs[1]->size(), indicators,

src/nbla/cuda/function/generic/inq_convolution.cu

@@ -150,8 +150,6 @@ void INQConvolutionCuda<T, T1>::setup_impl(const Variables &inputs,
     if (this->seed_ != -1) {
       // CURAND_RNG_PSEUDO_DEFAULT is CURAND_RNG_PSEUDO_XORWOW.
       curand_generator_ = curand_create_generator(this->seed_);
-    } else {
-      curand_generator_ = SingletonManager::get<Cuda>()->curand_generator();
     }
   }
 
@@ -252,7 +250,11 @@ void INQConvolutionCuda<T, T1>::forward_impl(const Variables &inputs,
       } else {
         // random selection (we re-use old_weights here to keep the random
         // values)
-        curand_generate_rand<Tc>(curand_generator_, 0.0f, 1.0f, old_weights,
+        curandGenerator_t &gen =
+            this->seed_ == -1
+                ? SingletonManager::get<Cuda>()->curand_generator()
+                : curand_generator_;
+        curand_generate_rand<Tc>(gen, 0.0f, 1.0f, old_weights,
                                  inputs[0]->size());
         NBLA_CUDA_LAUNCH_KERNEL_SIMPLE((kernel_random_selection<Tc, T1>),
                                        inputs[1]->size(), indicators,

src/nbla/cuda/function/generic/rand.cu

@@ -32,8 +32,11 @@ void RandCuda<T>::forward_impl(const Variables &inputs,
   // In any type, this uses float32 type.
   typedef typename CudaTypeForceFloat<T>::type Tc;
   cuda_set_device(device_);
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
   Tc *y = outputs[0]->cast_data_and_get_pointer<Tc>(this->ctx_, true);
-  curand_generate_rand<float>(curand_generator_, this->low_, this->high_, y,
+  curand_generate_rand<float>(gen, this->low_, this->high_, y,
                               outputs[0]->size());
 }
 

src/nbla/cuda/function/generic/randint.cu

@@ -30,8 +30,11 @@ template <typename T>
 void RandintCuda<T>::forward_impl(const Variables &inputs,
                                   const Variables &outputs) {
   cuda_set_device(device_);
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
   int *y = outputs[0]->cast_data_and_get_pointer<int>(this->ctx_, true);
-  curand_generate_rand<int>(curand_generator_, this->low_, this->high_, y,
+  curand_generate_rand<int>(gen, this->low_, this->high_, y,
                             outputs[0]->size());
 }
 

src/nbla/cuda/function/generic/randn.cu

@@ -31,8 +31,11 @@ void RandnCuda<T>::forward_impl(const Variables &inputs,
                                 const Variables &outputs) {
   typedef typename CudaTypeForceFloat<T>::type Tc;
   cuda_set_device(device_);
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
   Tc *y = outputs[0]->cast_data_and_get_pointer<Tc>(this->ctx_, true);
-  curand_generate_randn<float>(curand_generator_, this->mu_, this->sigma_, y,
+  curand_generate_randn<float>(gen, this->mu_, this->sigma_, y,
                                outputs[0]->size());
 }
 

src/nbla/cuda/function/generic/random_choice.cu

@@ -131,7 +131,10 @@ void RandomChoiceCuda<T>::sample_with_replacement(const Variables &inputs,
       tmp1.cast(get_dtype<float>(), this->ctx_, true)->pointer<float>();
 
   // Generate random choices for each output sample point.
-  curand_generate_rand<float>(curand_generator_, 0, 1, u_vals, y->size());
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<float>(gen, 0, 1, u_vals, y->size());
 
   // Build cumulative sum of weights per population.
   for (int i = 0; i < this->outer_loop_; i++) {
@@ -178,7 +181,10 @@ void RandomChoiceCuda<T>::sample_without_replace(const Variables &inputs,
                  thrust::device_pointer_cast(w_data));
 
   // Generate random choices for each output sample point.
-  curand_generate_rand<float>(curand_generator_, 0, 1, u_vals, y->size());
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<float>(gen, 0, 1, u_vals, y->size());
 
   // We draw one sample per round (and population) and set the choosen weight
   // to zero, so each round decreases the number of non-zero weights.

src/nbla/cuda/function/generic/random_crop.cu

@@ -116,7 +116,10 @@ void RandomCropCuda<T>::forward_impl(const Variables &inputs,
   this->random_values_ = make_shared<CudaCachedArray>(
       this->size_ * this->shape_.size(), dtypes::INT, this->ctx_);
   int *random_values = this->random_values_->template pointer<int>();
-  curand_generate_rand<int>(curand_generator_, 0, 2 ^ 23, random_values,
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<int>(gen, 0, 2 ^ 23, random_values,
                             this->size_ * this->shape_.size());
   const int *shape_info_gpu =
       this->shape_info_buf_.get(dtypes::INT, this->ctx_)

src/nbla/cuda/function/generic/random_erase.cu

@@ -191,8 +191,11 @@ void RandomEraseCuda<T>::forward_impl(const Variables &inputs,
       this->random_coordinates_->cast(get_dtype<float>(), this->ctx_)
           ->template pointer<float>();
 
-  curand_generate_rand<float>(this->curand_generator_, 0.0f, 1.0f,
-                              random_coords, this->random_coordinates_->size());
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<float>(gen, 0.0f, 1.0f, random_coords,
+                              this->random_coordinates_->size());
 
   // Create 5 x N x B (x C), 5 is {prob, ye_start, xe_start, ye_end, xe_end}
   // inplace

src/nbla/cuda/function/generic/random_flip.cu

@@ -82,7 +82,10 @@ void RandomFlipCuda<T>::forward_impl(const Variables &inputs,
       Shape_t{static_cast<Size_t>(this->size_ * inputs[0]->ndim())}, true);
   int *flip_flags = this->flip_flags_.cast(dtypes::INT, this->ctx_, true)
                         ->template pointer<int>();
-  curand_generate_rand<int>(curand_generator_, 0, 255, flip_flags,
+  curandGenerator_t &gen =
+      this->seed_ == -1 ? SingletonManager::get<Cuda>()->curand_generator()
+                        : curand_generator_;
+  curand_generate_rand<int>(gen, 0, 255, flip_flags,
                             this->size_ * inputs[0]->ndim());
   const Tcu *x = inputs[0]->get_data_pointer<Tcu>(this->ctx_);
   Tcu *y = outputs[0]->cast_data_and_get_pointer<Tcu>(this->ctx_, true);

src/nbla/cuda/utils/random.cpp

@@ -15,6 +15,7 @@
 #include <nbla/cuda/array/cuda_array.hpp>
 #include <nbla/cuda/utils/random.hpp>
 #include <nbla/nd_array.hpp>
+#include <nbla/random_manager.hpp>
 
 #include <random>
 
@@ -24,7 +25,7 @@ curandGenerator_t curand_create_generator(int seed) {
   curandGenerator_t gen;
   NBLA_CURAND_CHECK(curandCreateGenerator(&gen, CURAND_RNG_PSEUDO_DEFAULT));
   if (seed == -1) {
-    seed = std::random_device()();
+    seed = SingletonManager::get<RandomManager>()->get_seed();
   }
   curand_set_seed(gen, seed);
   return gen;