Merge pull request #465 from pdziekan/curand_dbg

urand: use gpuErrchk for debugging

Author: pdziekan

Readme.md

@@ -115,5 +115,3 @@ Some CMake hints:
 
 - the output of commands executed by "make test" can be viewed with:
   $ less Testing/Temporary/LastTest.log
-
-

src/detail/gpu_assert.hpp

@@ -6,6 +6,8 @@
 #include <mutex>
 #include <condition_variable>
 
+#include <curand.h>
+
 // macro to check for cuda errors, taken from 
 // http://stackoverflow.com/questions/14038589/what-is-the-canonical-way-to-check-for-errors-using-the-cuda-runtime-api
 #define gpuErrchk(ans) { detail::gpuAssert((ans), __FILE__, __LINE__); }
@@ -16,6 +18,57 @@ namespace libcloudphxx
   {
     namespace detail
     {   
+      // https://stackoverflow.com/questions/40704147/built-in-function-to-get-curands-curandstatus-t-as-string
+      #ifdef CURAND_H_
+      // cuRAND API errors
+      static const char *curandGetErrorString(curandStatus_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";
+          }
+      
+          return "<unknown>";
+      }
+      #endif
+
       inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
       {   
          if (code != cudaSuccess) 
@@ -25,6 +78,15 @@ namespace libcloudphxx
          }
       }   
 
+      inline void gpuAssert(curandStatus_t code, const char *file, int line, bool abort=true)
+      {   
+         if (code != CURAND_STATUS_SUCCESS) 
+         {
+            fprintf(stderr,"GPUassert (curand): %s %s %d\n", curandGetErrorString(code), file, line);
+            if (abort) exit(code);
+         }
+      }   
+
       // max(1, n)
       inline int m1(int n) { return n == 0 ? 1 : n; }
 

src/detail/urand.hpp

@@ -100,40 +100,26 @@ namespace libcloudphxx
 
         rng(int seed)
         {
-          {
-            int status = curandCreateGenerator(&gen, CURAND_RNG_PSEUDO_MTGP32);
-            assert(status == CURAND_STATUS_SUCCESS /* && "curandCreateGenerator failed"*/);
-            _unused(status);
-          }
-          {
-            int status = curandSetPseudoRandomGeneratorSeed(gen, seed);
-            assert(status == CURAND_STATUS_SUCCESS /* && "curandSetPseudoRandomGeneratorSeed failed"*/);
-            _unused(status);
-          }
+          gpuErrchk(curandCreateGenerator(&gen, CURAND_RNG_PSEUDO_MTGP32));
+          gpuErrchk(curandSetPseudoRandomGeneratorSeed(gen, seed));
         }
 
         void reseed(int seed)
         {
-          int status = curandSetPseudoRandomGeneratorSeed(gen, seed);
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandSetPseudoRandomGeneratorSeed failed"*/);
-          _unused(status);
+          gpuErrchk(curandSetPseudoRandomGeneratorSeed(gen, seed));
         }
 
         ~rng()
         {
-          int status = curandDestroyGenerator(gen); 
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandDestroyGenerator failed"*/);
-          _unused(status);
+          gpuErrchk(curandDestroyGenerator(gen));
         }
 
         void generate_n(
           thrust_device::vector<float> &v, 
           const thrust_size_t n
         )
         {
-          int status = curandGenerateUniform(gen, thrust::raw_pointer_cast(v.data()), n); // (0,1] range
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandGenerateUniform failed"*/);
-          _unused(status);
+          gpuErrchk(curandGenerateUniform(gen, thrust::raw_pointer_cast(v.data()), n)); // (0,1] range
           // shift into the expected [0,1) range
           namespace arg = thrust::placeholders;
           thrust::transform(v.begin(), v.begin() + n, v.begin(), float(1) - arg::_1);
@@ -144,9 +130,7 @@ namespace libcloudphxx
           const thrust_size_t n
         )
         {
-          int status = curandGenerateUniformDouble(gen, thrust::raw_pointer_cast(v.data()), n); // (0,1] range
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandGenerateUniform failed"*/);
-          _unused(status);
+          gpuErrchk(curandGenerateUniformDouble(gen, thrust::raw_pointer_cast(v.data()), n)); // (0,1] range
           // shift into the expected [0,1) range
           namespace arg = thrust::placeholders;
           thrust::transform(v.begin(), v.begin() + n, v.begin(), double(1) - arg::_1);
@@ -157,29 +141,23 @@ namespace libcloudphxx
           const thrust_size_t n
         )
         {
-          int status = curandGenerateNormal(gen, thrust::raw_pointer_cast(v.data()), n, float(0), float(1));
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandGenerateUniform failed"*/);
-          _unused(status);
+          gpuErrchk(curandGenerateNormal(gen, thrust::raw_pointer_cast(v.data()), n, float(0), float(1)));
         }
 
         void generate_normal_n(
           thrust_device::vector<double> &v, 
           const thrust_size_t n
         )
         {
-          int status = curandGenerateNormalDouble(gen, thrust::raw_pointer_cast(v.data()), n, double(0), double(1));
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandGenerateUniform failed"*/);
-          _unused(status);
+          gpuErrchk(curandGenerateNormalDouble(gen, thrust::raw_pointer_cast(v.data()), n, double(0), double(1)));
         }
 
         void generate_n(
           thrust_device::vector<unsigned int> &v, 
           const thrust_size_t n
         )
         {
-          int status = curandGenerate(gen, thrust::raw_pointer_cast(v.data()), n);
-          assert(status == CURAND_STATUS_SUCCESS /* && "curandGenerateUniform failed"*/);
-          _unused(status);
+          gpuErrchk(curandGenerate(gen, thrust::raw_pointer_cast(v.data()), n));
         }
 #endif
       };