3 files changed, 8 insertions, 10 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
     }
 
diff --git a/nerv/lib/matrix/generic/cumatrix.h b/nerv/lib/matrix/generic/cumatrix.h
index e82dccd..9304060 100644
--- a/nerv/lib/matrix/generic/cumatrix.h
+++ b/nerv/lib/matrix/generic/cumatrix.h
@@ -61,3 +61,7 @@ void nerv_matrix_(scale_rows_by_col)(Matrix *a, const Matrix *b,
                                     Status *status);
 void nerv_matrix_(scale_rows_by_row)(Matrix *a, const Matrix *b,
                                     Status *status);
+void nerv_matrix_(thres_mask)(Matrix *a, Matrix *b,
+                                double thres, double low, double high,
+                                Status *status);
+void nerv_matrix_(rand_uniform)(Matrix *a, Status *status);
diff --git a/nerv/lib/matrix/generic/matrix.c b/nerv/lib/matrix/generic/matrix.c
index 004d9aa..998d107 100644
--- a/nerv/lib/matrix/generic/matrix.c
+++ b/nerv/lib/matrix/generic/matrix.c
@@ -9,8 +9,6 @@ void nerv_matrix_(data_free)(Matrix *self, Status *status) {
     {
         /* free matrix data */
         MATRIX_DATA_FREE(MATRIX_ELEM_PTR_BASE(self), status);
-        curandDestroyGenerator(*(self->curand_gen));
-        free(self->curand_gen);
         free(self->data_ref);
         free(self);
     }
@@ -40,11 +38,6 @@ Matrix *nerv_matrix_(create)(long nrow, long ncol, Status *status) {
     }
     self->data_ref = (long *)malloc(sizeof(long));
     *self->data_ref = 0;
-    
-    self->curand_gen = (curandGenerator_t*)malloc(sizeof(curandGenerator_t));
-    curandCreateGenerator(self->curand_gen, CURAND_RNG_PSEUDO_DEFAULT); 
-    curandSetPseudoRandomGeneratorSeed(*(self->curand_gen), time(NULL));
-
     self->offset = 0;
     nerv_matrix_(data_retain)(self);
     NERV_SET_STATUS(status, NERV_NORMAL, 0);
@@ -64,7 +57,6 @@ Matrix *nerv_matrix_(getrow)(Matrix *self, int row) {
     prow->nmax = prow->ncol;
     prow->data = self->data;
     prow->data_ref = self->data_ref;
-    prow->curand_gen = self->curand_gen;
     prow->offset = row * self->stride;
     nerv_matrix_(data_retain)(prow);
     return prow;