diff options
| author | Determinant <ted.sybil@gmail.com> | 2016-02-15 16:18:13 +0800 |
|---|---|---|
| committer | Determinant <ted.sybil@gmail.com> | 2016-02-15 16:18:13 +0800 |
| commit | a7608a67f16f745309409f6a004354121e0b3ba6 (patch) | |
| tree | b36ec7c4e8b252f540f049d374a13915ca5ba9a3 /nerv/lib/matrix/generic/cukernel.cu | |
| parent | 3362020a6bc43766a92882abe6d127c8bb98a628 (diff) | |
move curand_gen out of the matrix layout
Diffstat (limited to 'nerv/lib/matrix/generic/cukernel.cu')
| -rw-r--r-- | nerv/lib/matrix/generic/cukernel.cu | 6 |
1 files changed, 4 insertions, 2 deletions
diff --git a/nerv/lib/matrix/generic/cukernel.cu b/nerv/lib/matrix/generic/cukernel.cu index 7f780a8..8fbe05d 100644 --- a/nerv/lib/matrix/generic/cukernel.cu +++ b/nerv/lib/matrix/generic/cukernel.cu @@ -4,6 +4,7 @@ #include "../matrix.h" #include "cuda.h" #include "float.h" +#include "curand.h" #define CUDA_THREADS_N 16 #define CUDA_THREADS_NN ((CUDA_THREADS_N) * (CUDA_THREADS_N)) #define CEIL_DIV(a, b) (((a) + (b) - 1) / (b)) @@ -430,12 +431,13 @@ extern "C" { cudaStreamSynchronize(0); } + extern curandGenerator_t curand_gen; void cudak_(cuda_rand_uniform)(const Matrix *a) { #ifdef MATRIX_USE_FLOAT - curandGenerateUniform(*(a->curand_gen), MATRIX_ELEM_PTR(a), a->nrow * a->stride / sizeof(MATRIX_ELEM)); + curandGenerateUniform(curand_gen, MATRIX_ELEM_PTR(a), a->nrow * a->stride / sizeof(MATRIX_ELEM)); #endif #ifdef MATRIX_USE_DOUBLE - curandGenerateUniformDouble(*(a->curand_gen), MATRIX_ELEM_PTR(a), a->nrow * a->stride / sizeof(MATRIX_ELEM)); + curandGenerateUniformDouble(curand_gen, MATRIX_ELEM_PTR(a), a->nrow * a->stride / sizeof(MATRIX_ELEM)); #endif } |