luke
/
ngspice
mirror of https://git.code.sf.net/p/ngspice/ngspice
1
0
Fork 0
Code Issues Packages Projects Releases Wiki Activity

Updated the CUSPICE RES model

This commit is contained in:
Francesco Lannutti 2018-04-23 22:56:59 +02:00
parent df106f2c0a
commit 82b62546f9
6 changed files with 14 additions and 144 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

27
src/spicelib/devices/res/CUSPICE/curesfree.c
View File

@ -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) ;

21
src/spicelib/devices/res/CUSPICE/curesload.cu
View File

@ -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] ;
}
}

36
src/spicelib/devices/res/CUSPICE/curessetup.c
View File

@ -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)

24
src/spicelib/devices/res/CUSPICE/curestemp.c
View File

@ -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)

42
src/spicelib/devices/res/resdefs.h
View File

@ -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

8
src/spicelib/devices/res/restemp.c
View File

@ -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