diff --git a/src/spicelib/devices/res/CUSPICE/curesfree.c b/src/spicelib/devices/res/CUSPICE/curesfree.c index 6efa9fbc7..d3488907d 100644 --- a/src/spicelib/devices/res/CUSPICE/curesfree.c +++ b/src/spicelib/devices/res/CUSPICE/curesfree.c @@ -39,40 +39,13 @@ GENmodel *inModel for ( ; model != NULL ; model = RESnextModel(model)) { /* DOUBLE */ - free (model->RESparamCPU.REStc1Array) ; - cudaFree (model->RESparamGPU.d_REStc1Array) ; - - free (model->RESparamCPU.REStc2Array) ; - cudaFree (model->RESparamGPU.d_REStc2Array) ; - - free (model->RESparamCPU.RESmArray) ; - cudaFree (model->RESparamGPU.d_RESmArray) ; - free (model->RESparamCPU.RESconductArray) ; cudaFree (model->RESparamGPU.d_RESconductArray) ; - free (model->RESparamCPU.REStempArray) ; - cudaFree (model->RESparamGPU.d_REStempArray) ; - - free (model->RESparamCPU.RESdtempArray) ; - cudaFree (model->RESparamGPU.d_RESdtempArray) ; - free (model->RESparamCPU.REScurrentArray) ; cudaFree (model->RESparamGPU.d_REScurrentArray) ; - free (model->RESparamCPU.RESgValueArray) ; - cudaFree (model->RESparamGPU.d_RESgValueArray) ; - /* INT */ - free (model->RESparamCPU.REStc1GivenArray) ; - cudaFree (model->RESparamGPU.d_REStc1GivenArray) ; - - free (model->RESparamCPU.REStc2GivenArray) ; - cudaFree (model->RESparamGPU.d_REStc2GivenArray) ; - - free (model->RESparamCPU.RESmGivenArray) ; - cudaFree (model->RESparamGPU.d_RESmGivenArray) ; - free (model->RESparamCPU.RESposNodeArray) ; cudaFree (model->RESparamGPU.d_RESposNodeArray) ; diff --git a/src/spicelib/devices/res/CUSPICE/curesload.cu b/src/spicelib/devices/res/CUSPICE/curesload.cu index ebeac3a4c..1170d3338 100644 --- a/src/spicelib/devices/res/CUSPICE/curesload.cu +++ b/src/spicelib/devices/res/CUSPICE/curesload.cu @@ -81,8 +81,6 @@ cuRESload_kernel RESparamGPUstruct RESentry, double *CKTrhsOld, int n_instances, int *d_PositionVector, double * d_CKTloadOutput ) { - double m, difference, factor ; - int instance_ID ; instance_ID = threadIdx.y + blockDim.y * blockIdx.x ; @@ -90,26 +88,11 @@ RESparamGPUstruct RESentry, double *CKTrhsOld, int n_instances, int *d_PositionV { if (threadIdx.x == 0) { - if (!(RESentry.d_REStc1GivenArray [instance_ID])) - RESentry.d_REStc1Array [instance_ID] = 0.0 ; - - if (!(RESentry.d_REStc2GivenArray [instance_ID])) - RESentry.d_REStc2Array [instance_ID] = 0.0 ; - - if (!(RESentry.d_RESmGivenArray [instance_ID])) - RESentry.d_RESmArray [instance_ID] = 1.0 ; - RESentry.d_REScurrentArray [instance_ID] = (CKTrhsOld [RESentry.d_RESposNodeArray [instance_ID]] - CKTrhsOld [RESentry.d_RESnegNodeArray [instance_ID]]) * RESentry.d_RESconductArray [instance_ID] ; - - difference = (RESentry.d_REStempArray [instance_ID] + RESentry.d_RESdtempArray [instance_ID]) - 300.15 ; - factor = 1.0 + (RESentry.d_REStc1Array [instance_ID]) * difference + - (RESentry.d_REStc2Array [instance_ID]) * difference * difference ; - - m = (RESentry.d_RESmArray [instance_ID]) / factor ; - - d_CKTloadOutput [d_PositionVector [instance_ID]] = m * RESentry.d_RESconductArray [instance_ID] ; + + d_CKTloadOutput [d_PositionVector [instance_ID]] = RESentry.d_RESconductArray [instance_ID] ; } } diff --git a/src/spicelib/devices/res/CUSPICE/curessetup.c b/src/spicelib/devices/res/CUSPICE/curessetup.c index 444389bad..ddcebf5cb 100644 --- a/src/spicelib/devices/res/CUSPICE/curessetup.c +++ b/src/spicelib/devices/res/CUSPICE/curessetup.c @@ -68,51 +68,15 @@ GENmodel *inModel CUDAMEMCPYCHECK (model->d_PositionVector, size, int, status) /* DOUBLE */ - model->RESparamCPU.REStc1Array = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_REStc1Array), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_REStc1Array, size, double, status) - - model->RESparamCPU.REStc2Array = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_REStc2Array), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_REStc2Array, size, double, status) - - model->RESparamCPU.RESmArray = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESmArray), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_RESmArray, size, double, status) - model->RESparamCPU.RESconductArray = (double *) malloc (size * sizeof(double)) ; status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESconductArray), size * sizeof(double)) ; CUDAMALLOCCHECK (model->RESparamGPU.d_RESconductArray, size, double, status) - model->RESparamCPU.REStempArray = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_REStempArray), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_REStempArray, size, double, status) - - model->RESparamCPU.RESdtempArray = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESdtempArray), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_RESdtempArray, size, double, status) - model->RESparamCPU.REScurrentArray = (double *) malloc (size * sizeof(double)) ; status = cudaMalloc ((void **)&(model->RESparamGPU.d_REScurrentArray), size * sizeof(double)) ; CUDAMALLOCCHECK (model->RESparamGPU.d_REScurrentArray, size, double, status) - model->RESparamCPU.RESgValueArray = (double *) malloc (size * sizeof(double)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESgValueArray), size * sizeof(double)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_RESgValueArray, size, double, status) - /* INT */ - model->RESparamCPU.REStc1GivenArray = (int *) malloc (size * sizeof(int)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_REStc1GivenArray), size * sizeof(int)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_REStc1GivenArray, size, int, status) - - model->RESparamCPU.REStc2GivenArray = (int *) malloc (size * sizeof(int)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_REStc2GivenArray), size * sizeof(int)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_REStc2GivenArray, size, int, status) - - model->RESparamCPU.RESmGivenArray = (int *) malloc (size * sizeof(int)) ; - status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESmGivenArray), size * sizeof(int)) ; - CUDAMALLOCCHECK (model->RESparamGPU.d_RESmGivenArray, size, int, status) - model->RESparamCPU.RESposNodeArray = (int *) malloc (size * sizeof(int)) ; status = cudaMalloc ((void **)&(model->RESparamGPU.d_RESposNodeArray), size * sizeof(int)) ; CUDAMALLOCCHECK (model->RESparamGPU.d_RESposNodeArray, size, int, status) diff --git a/src/spicelib/devices/res/CUSPICE/curestemp.c b/src/spicelib/devices/res/CUSPICE/curestemp.c index b996764ee..957cf732a 100644 --- a/src/spicelib/devices/res/CUSPICE/curestemp.c +++ b/src/spicelib/devices/res/CUSPICE/curestemp.c @@ -51,34 +51,10 @@ GENmodel *inModel size = (long unsigned int) model->n_instances; /* DOUBLE */ - status = cudaMemcpy (model->RESparamGPU.d_REStc1Array, model->RESparamCPU.REStc1Array, size * sizeof(double), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_REStc1Array, size, double, status) - - status = cudaMemcpy (model->RESparamGPU.d_REStc2Array, model->RESparamCPU.REStc2Array, size * sizeof(double), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_REStc2Array, size, double, status) - - status = cudaMemcpy (model->RESparamGPU.d_RESmArray, model->RESparamCPU.RESmArray, size * sizeof(double), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_RESmArray, size, double, status) - status = cudaMemcpy (model->RESparamGPU.d_RESconductArray, model->RESparamCPU.RESconductArray, size * sizeof(double), cudaMemcpyHostToDevice) ; CUDAMEMCPYCHECK(model->RESparamGPU.d_RESconductArray, size, double, status) - status = cudaMemcpy (model->RESparamGPU.d_REStempArray, model->RESparamCPU.REStempArray, size * sizeof(double), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_REStempArray, size, double, status) - - status = cudaMemcpy (model->RESparamGPU.d_RESdtempArray, model->RESparamCPU.RESdtempArray, size * sizeof(double), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_RESdtempArray, size, double, status) - /* INT */ - status = cudaMemcpy (model->RESparamGPU.d_REStc1GivenArray, model->RESparamCPU.REStc1GivenArray, size * sizeof(int), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_REStc1GivenArray, size, int, status) - - status = cudaMemcpy (model->RESparamGPU.d_REStc2GivenArray, model->RESparamCPU.REStc2GivenArray, size * sizeof(int), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_REStc2GivenArray, size, int, status) - - status = cudaMemcpy (model->RESparamGPU.d_RESmGivenArray, model->RESparamCPU.RESmGivenArray, size * sizeof(int), cudaMemcpyHostToDevice) ; - CUDAMEMCPYCHECK(model->RESparamGPU.d_RESmGivenArray, size, int, status) - status = cudaMemcpy (model->RESparamGPU.d_RESposNodeArray, model->RESparamCPU.RESposNodeArray, size * sizeof(int), cudaMemcpyHostToDevice) ; CUDAMEMCPYCHECK(model->RESparamGPU.d_RESposNodeArray, size, int, status) diff --git a/src/spicelib/devices/res/resdefs.h b/src/spicelib/devices/res/resdefs.h index 716ff8522..6d394f8d2 100644 --- a/src/spicelib/devices/res/resdefs.h +++ b/src/spicelib/devices/res/resdefs.h @@ -97,41 +97,23 @@ typedef struct sRESinstance { #ifdef USE_CUSPICE typedef struct sRESparamCPUstruct { - double *REScpuPointersD [8] ; - #define REStc1Array REScpuPointersD[0] - #define REStc2Array REScpuPointersD[1] - #define RESmArray REScpuPointersD[2] - #define RESconductArray REScpuPointersD[3] - #define REStempArray REScpuPointersD[4] - #define RESdtempArray REScpuPointersD[5] - #define REScurrentArray REScpuPointersD[6] - #define RESgValueArray REScpuPointersD[7] + double *REScpuPointersD [2] ; + #define RESconductArray REScpuPointersD[0] + #define REScurrentArray REScpuPointersD[1] - int *REScpuPointersI [5] ; - #define REStc1GivenArray REScpuPointersI[0] - #define REStc2GivenArray REScpuPointersI[1] - #define RESmGivenArray REScpuPointersI[2] - #define RESposNodeArray REScpuPointersI[3] - #define RESnegNodeArray REScpuPointersI[4] + int *REScpuPointersI [2] ; + #define RESposNodeArray REScpuPointersI[0] + #define RESnegNodeArray REScpuPointersI[1] } RESparamCPUstruct ; typedef struct sRESparamGPUstruct { - double *REScudaPointersD [8] ; - #define d_REStc1Array REScudaPointersD[0] - #define d_REStc2Array REScudaPointersD[1] - #define d_RESmArray REScudaPointersD[2] - #define d_RESconductArray REScudaPointersD[3] - #define d_REStempArray REScudaPointersD[4] - #define d_RESdtempArray REScudaPointersD[5] - #define d_REScurrentArray REScudaPointersD[6] - #define d_RESgValueArray REScudaPointersD[7] + double *REScudaPointersD [2] ; + #define d_RESconductArray REScudaPointersD[0] + #define d_REScurrentArray REScudaPointersD[1] - int *REScudaPointersI [5] ; - #define d_REStc1GivenArray REScudaPointersI[0] - #define d_REStc2GivenArray REScudaPointersI[1] - #define d_RESmGivenArray REScudaPointersI[2] - #define d_RESposNodeArray REScudaPointersI[3] - #define d_RESnegNodeArray REScudaPointersI[4] + int *REScudaPointersI [2] ; + #define d_RESposNodeArray REScudaPointersI[0] + #define d_RESnegNodeArray REScudaPointersI[1] } RESparamGPUstruct ; #endif diff --git a/src/spicelib/devices/res/restemp.c b/src/spicelib/devices/res/restemp.c index cf1cc3120..57656a77e 100644 --- a/src/spicelib/devices/res/restemp.c +++ b/src/spicelib/devices/res/restemp.c @@ -53,17 +53,9 @@ REStemp(GENmodel *inModel, CKTcircuit *ckt) RESupdate_conduct(here, TRUE); #ifdef USE_CUSPICE - model->RESparamCPU.REStc1GivenArray[i] = here->REStc1Given; - model->RESparamCPU.REStc2GivenArray[i] = here->REStc2Given; - model->RESparamCPU.RESmGivenArray[i] = here->RESmGiven; - model->RESparamCPU.REStc1Array[i] = here->REStc1; - model->RESparamCPU.REStc2Array[i] = here->REStc2; - model->RESparamCPU.RESmArray[i] = here->RESm; model->RESparamCPU.RESposNodeArray[i] = here->RESposNode; model->RESparamCPU.RESnegNodeArray[i] = here->RESnegNode; model->RESparamCPU.RESconductArray[i] = here->RESconduct; - model->RESparamCPU.REStempArray[i] = here->REStemp; - model->RESparamCPU.RESdtempArray[i] = here->RESdtemp; i++; #endif