From 1b0f2ab768e34f126ce946e1689cd47c86f41645 Mon Sep 17 00:00:00 2001 From: Determinant Date: Thu, 10 Dec 2015 21:46:11 +0800 Subject: fix the bug in freeing submatrix; add `offset` to Matrix structure --- nerv/lib/matrix/cukernel.cu | 1 + nerv/lib/matrix/cumatrix.c | 1 + nerv/lib/matrix/generic/cumatrix.c | 2 +- nerv/lib/matrix/generic/elem_type.h | 13 ++++++++++--- nerv/lib/matrix/generic/matrix.c | 9 +++++---- nerv/lib/matrix/matrix.h | 1 + nerv/lib/matrix/mmatrix.c | 5 ++++- nerv/matrix/cumatrix.c | 1 + nerv/matrix/generic/cumatrix.c | 2 +- nerv/matrix/generic/elem_type.h | 22 ---------------------- nerv/matrix/generic/mmatrix.c | 2 +- nerv/matrix/mmatrix.c | 2 ++ 12 files changed, 28 insertions(+), 33 deletions(-) delete mode 100644 nerv/matrix/generic/elem_type.h diff --git a/nerv/lib/matrix/cukernel.cu b/nerv/lib/matrix/cukernel.cu index 210e6bf..c20e538 100644 --- a/nerv/lib/matrix/cukernel.cu +++ b/nerv/lib/matrix/cukernel.cu @@ -44,6 +44,7 @@ __device__ float atomicAdd_nvidia(float* address, float val) { #undef MATRIX_USE_FLOAT #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT diff --git a/nerv/lib/matrix/cumatrix.c b/nerv/lib/matrix/cumatrix.c index ff1168d..a8ed075 100644 --- a/nerv/lib/matrix/cumatrix.c +++ b/nerv/lib/matrix/cumatrix.c @@ -57,6 +57,7 @@ void nerv_cumatrix_init() { #undef MATRIX_USE_FLOAT #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT #undef MATRIX_CUMATRIX_HOST_TNAME diff --git a/nerv/lib/matrix/generic/cumatrix.c b/nerv/lib/matrix/generic/cumatrix.c index 00af895..7643c01 100644 --- a/nerv/lib/matrix/generic/cumatrix.c +++ b/nerv/lib/matrix/generic/cumatrix.c @@ -315,7 +315,7 @@ void nerv_matrix_(copy_rows_fromh_by_idx)(Matrix *a, const Matrix *b, long nrow = a->nrow; if (!(0 <= b_begin && b_begin + nrow <= idx->ncol)) NERV_EXIT_STATUS(status, MAT_INVALID_COPY_INTERVAL, 0); - float *idx_ptr = idx->data.f; + float *idx_ptr = MATRIX_ELEM_PTR_F(idx); int i; if (idx->nrow != 1) NERV_EXIT_STATUS(status, MAT_IDX_VECTOR_EXP, 0); diff --git a/nerv/lib/matrix/generic/elem_type.h b/nerv/lib/matrix/generic/elem_type.h index bffe940..07f6355 100644 --- a/nerv/lib/matrix/generic/elem_type.h +++ b/nerv/lib/matrix/generic/elem_type.h @@ -1,22 +1,29 @@ +#define MATRIX_ELEM_PTR_F(self) ((float *)((char *)((self)->data.f) + (self)->offset)) +#define MATRIX_ELEM_PTR_D(self) ((double *)((char *)((self)->data.d) + (self)->offset)) +#define MATRIX_ELEM_PTR_I(self) ((long *)((char *)((self)->data.i) + (self)->offset)) + #ifdef MATRIX_USE_FLOAT #define MATRIX_ELEM float #define MATRIX_ELEM_FMT "%f" #define MATRIX_ELEM_WRITE_FMT "%.8f" -#define MATRIX_ELEM_PTR(self) ((self)->data.f) +#define MATRIX_ELEM_PTR(self) MATRIX_ELEM_PTR_F(self) +#define MATRIX_ELEM_PTR_BASE(self) ((self)->data.f) #elif defined(MATRIX_USE_DOUBLE) #define MATRIX_ELEM double #define MATRIX_ELEM_FMT "%lf" #define MATRIX_ELEM_WRITE_FMT "%.8lf" -#define MATRIX_ELEM_PTR(self) ((self)->data.d) +#define MATRIX_ELEM_PTR(self) MATRIX_ELEM_PTR_D(self) +#define MATRIX_ELEM_PTR_BASE(self) ((self)->data.d) #elif defined(MATRIX_USE_INT) #define MATRIX_ELEM long #define MATRIX_ELEM_FMT "%ld" #define MATRIX_ELEM_WRITE_FMT "%ld" -#define MATRIX_ELEM_PTR(self) ((self)->data.i) +#define MATRIX_ELEM_PTR(self) MATRIX_ELEM_PTR_I(self) +#define MATRIX_ELEM_PTR_BASE(self) ((self)->data.i) #endif diff --git a/nerv/lib/matrix/generic/matrix.c b/nerv/lib/matrix/generic/matrix.c index 4246751..998d107 100644 --- a/nerv/lib/matrix/generic/matrix.c +++ b/nerv/lib/matrix/generic/matrix.c @@ -4,12 +4,11 @@ /* FIXME: malloc failure detection */ void nerv_matrix_(data_free)(Matrix *self, Status *status) { - if(*self->data_ref == 0) return; /* FIXME: repeat free memory */ assert(*self->data_ref > 0); if (--(*self->data_ref) == 0) { /* free matrix data */ - MATRIX_DATA_FREE(MATRIX_ELEM_PTR(self), status); + MATRIX_DATA_FREE(MATRIX_ELEM_PTR_BASE(self), status); free(self->data_ref); free(self); } @@ -29,7 +28,7 @@ Matrix *nerv_matrix_(create)(long nrow, long ncol, Status *status) { self->ncol = ncol; self->nmax = self->nrow * self->ncol; self->dim = 2; - MATRIX_DATA_ALLOC(&MATRIX_ELEM_PTR(self), &self->stride, + MATRIX_DATA_ALLOC(&MATRIX_ELEM_PTR_BASE(self), &self->stride, sizeof(MATRIX_ELEM) * self->ncol, self->nrow, status); if (status->err_code != NERV_NORMAL) @@ -39,6 +38,7 @@ Matrix *nerv_matrix_(create)(long nrow, long ncol, Status *status) { } self->data_ref = (long *)malloc(sizeof(long)); *self->data_ref = 0; + self->offset = 0; nerv_matrix_(data_retain)(self); NERV_SET_STATUS(status, NERV_NORMAL, 0); return self; @@ -55,8 +55,9 @@ Matrix *nerv_matrix_(getrow)(Matrix *self, int row) { prow->dim = 1; prow->stride = self->stride; prow->nmax = prow->ncol; - MATRIX_ELEM_PTR(prow) = MATRIX_ROW_PTR(self, row); + prow->data = self->data; prow->data_ref = self->data_ref; + prow->offset = row * self->stride; nerv_matrix_(data_retain)(prow); return prow; } diff --git a/nerv/lib/matrix/matrix.h b/nerv/lib/matrix/matrix.h index 67a6e30..073bd13 100644 --- a/nerv/lib/matrix/matrix.h +++ b/nerv/lib/matrix/matrix.h @@ -12,6 +12,7 @@ typedef struct Matrix { double *d; long *i; } data; /* pointer to actual storage */ + unsigned long offset; /* the actual beginning of the matrix */ long *data_ref; } Matrix; diff --git a/nerv/lib/matrix/mmatrix.c b/nerv/lib/matrix/mmatrix.c index b8157eb..b5670f2 100644 --- a/nerv/lib/matrix/mmatrix.c +++ b/nerv/lib/matrix/mmatrix.c @@ -6,6 +6,7 @@ #define host_matrix_(NAME) host_matrix_float_##NAME #define nerv_matrix_(NAME) nerv_matrix_host_float_##NAME #include "generic/matrix.h" +#include "generic/elem_type.h" #include "generic/mmatrix.c" Matrix *nerv_matrix_(perm_gen)(int ncol, Status *status) { @@ -13,7 +14,7 @@ Matrix *nerv_matrix_(perm_gen)(int ncol, Status *status) { Matrix *self = nerv_matrix_(create)(1, ncol, status); if (status->err_code != NERV_NORMAL) return NULL; - float *prow = self->data.f; + float *prow = MATRIX_ELEM_PTR_F(self); for (i = 0; i < ncol; i++) prow[i] = i; for (i = ncol - 1; i >= 0; i--) @@ -31,6 +32,7 @@ Matrix *nerv_matrix_(perm_gen)(int ncol, Status *status) { #undef MATRIX_USE_FLOAT #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT @@ -44,6 +46,7 @@ Matrix *nerv_matrix_(perm_gen)(int ncol, Status *status) { #undef MATRIX_USE_DOUBLE #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT diff --git a/nerv/matrix/cumatrix.c b/nerv/matrix/cumatrix.c index fef03fc..bf92f92 100644 --- a/nerv/matrix/cumatrix.c +++ b/nerv/matrix/cumatrix.c @@ -49,6 +49,7 @@ const char *nerv_matrix_(tname) = "nerv.CuMatrixFloat"; #undef MATRIX_USE_FLOAT #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT #undef MATRIX_CUMATRIX_HOST_TNAME diff --git a/nerv/matrix/generic/cumatrix.c b/nerv/matrix/generic/cumatrix.c index e1519b0..be3d627 100644 --- a/nerv/matrix/generic/cumatrix.c +++ b/nerv/matrix/generic/cumatrix.c @@ -1,5 +1,5 @@ #ifdef NERV_GENERIC_CUMATRIX -#include "elem_type.h" +#include "../../lib/matrix/generic/elem_type.h" #define MATRIX_DATA_WRITE(L, data, idx, val) cuda_matrix_(write)(L, data, idx, val) #define MATRIX_DATA_READ(L, data, idx) cuda_matrix_(read)(L, data, idx) #define MATRIX_INIT(L) cuda_matrix_(init)(L) diff --git a/nerv/matrix/generic/elem_type.h b/nerv/matrix/generic/elem_type.h deleted file mode 100644 index bffe940..0000000 --- a/nerv/matrix/generic/elem_type.h +++ /dev/null @@ -1,22 +0,0 @@ -#ifdef MATRIX_USE_FLOAT - -#define MATRIX_ELEM float -#define MATRIX_ELEM_FMT "%f" -#define MATRIX_ELEM_WRITE_FMT "%.8f" -#define MATRIX_ELEM_PTR(self) ((self)->data.f) - -#elif defined(MATRIX_USE_DOUBLE) - -#define MATRIX_ELEM double -#define MATRIX_ELEM_FMT "%lf" -#define MATRIX_ELEM_WRITE_FMT "%.8lf" -#define MATRIX_ELEM_PTR(self) ((self)->data.d) - -#elif defined(MATRIX_USE_INT) - -#define MATRIX_ELEM long -#define MATRIX_ELEM_FMT "%ld" -#define MATRIX_ELEM_WRITE_FMT "%ld" -#define MATRIX_ELEM_PTR(self) ((self)->data.i) - -#endif diff --git a/nerv/matrix/generic/mmatrix.c b/nerv/matrix/generic/mmatrix.c index 01dd9e5..a4e8489 100644 --- a/nerv/matrix/generic/mmatrix.c +++ b/nerv/matrix/generic/mmatrix.c @@ -1,6 +1,6 @@ #ifdef NERV_GENERIC_MMATRIX #include "../../lib/matrix/generic/matrix.h" -#include "elem_type.h" +#include "../../lib/matrix/generic/elem_type.h" #define MATRIX_DATA_WRITE(L, data, idx, val) (data[idx] = val) #define MATRIX_DATA_READ(L, data, idx) (data[idx]) #define MATRIX_INIT(L) host_matrix_(init)(L) diff --git a/nerv/matrix/mmatrix.c b/nerv/matrix/mmatrix.c index 961059c..20c31d6 100644 --- a/nerv/matrix/mmatrix.c +++ b/nerv/matrix/mmatrix.c @@ -45,6 +45,7 @@ static const luaL_Reg nerv_matrix_(extra_methods_int)[] = { #undef MATRIX_USE_FLOAT #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT #undef MMATRIX_INIT @@ -60,6 +61,7 @@ const char *nerv_matrix_(tname) = "nerv.MMatrixDouble"; #undef MATRIX_USE_DOUBLE #undef MATRIX_ELEM #undef MATRIX_ELEM_PTR +#undef MATRIX_ELEM_PTR_BASE #undef MATRIX_ELEM_FMT #undef MATRIX_ELEM_WRITE_FMT -- cgit v1.2.3-70-g09d2