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

whitespace, indentation for the next commit

This commit is contained in:
Francesco Lannutti 2017-09-19 21:19:23 +02:00 committed by rlar
parent ebfc69f1e0
commit a4a3da9dfd
9 changed files with 213 additions and 213 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

14
src/spicelib/analysis/CUSPICE/cucktflush.c
View File

@ -35,16 +35,16 @@ cuCKTflush
CKTcircuit *ckt CKTcircuit *ckt
) )
{ {
long unsigned int m, mRHS ; long unsigned int m, mRHS ;
m = (long unsigned int)(ckt->total_n_values + 1) ; // + 1 because of CKTdiagGmin m = (long unsigned int)(ckt->total_n_values + 1) ; // + 1 because of CKTdiagGmin
mRHS = (long unsigned int)ckt->total_n_valuesRHS ; mRHS = (long unsigned int)ckt->total_n_valuesRHS ;
/* Clean-up the CKTloadOutput */ /* Clean-up the CKTloadOutput */
cudaMemset (ckt->d_CKTloadOutput, 0, m * sizeof(double)) ; cudaMemset (ckt->d_CKTloadOutput, 0, m * sizeof(double)) ;
/* Clean-up the CKTloadOutputRHS */ /* Clean-up the CKTloadOutputRHS */
cudaMemset (ckt->d_CKTloadOutputRHS, 0, mRHS * sizeof(double)) ; cudaMemset (ckt->d_CKTloadOutputRHS, 0, mRHS * sizeof(double)) ;
return (OK) ; return (OK) ;
} }

8
src/spicelib/analysis/CUSPICE/cucktnonconupdate.c
View File

@ -47,8 +47,8 @@ CKTcircuit *ckt
{ {
cudaError_t status ; cudaError_t status ;
status = cudaMemcpy (ckt->d_CKTnoncon, &(ckt->CKTnoncon), sizeof(int), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTnoncon, &(ckt->CKTnoncon), sizeof(int), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTnoncon, 1, int, status) CUDAMEMCPYCHECK (ckt->d_CKTnoncon, 1, int, status)
return (OK) ; return (OK) ;
} }
@ -61,8 +61,8 @@ CKTcircuit *ckt
{ {
cudaError_t status ; cudaError_t status ;
status = cudaMemcpy (&(ckt->CKTnoncon), ckt->d_CKTnoncon, sizeof(int), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (&(ckt->CKTnoncon), ckt->d_CKTnoncon, sizeof(int), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (&(ckt->CKTnoncon), 1, int, status) CUDAMEMCPYCHECK (&(ckt->CKTnoncon), 1, int, status)
return (OK) ; return (OK) ;
} }

16
src/spicelib/analysis/CUSPICE/cucktrhsoldupdate.c
View File

@ -47,8 +47,8 @@ CKTcircuit *ckt
{ {
long unsigned int size ; long unsigned int size ;
size = (long unsigned int)(ckt->d_MatrixSize + 1) ; size = (long unsigned int)(ckt->d_MatrixSize + 1) ;
cudaMemset (ckt->d_CKTrhsOld, 0, size * sizeof(double)) ; cudaMemset (ckt->d_CKTrhsOld, 0, size * sizeof(double)) ;
return (OK) ; return (OK) ;
} }
@ -62,9 +62,9 @@ CKTcircuit *ckt
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)(ckt->d_MatrixSize + 1) ; size = (long unsigned int)(ckt->d_MatrixSize + 1) ;
status = cudaMemcpy (ckt->d_CKTrhsOld, ckt->CKTrhsOld, size * sizeof(double), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTrhsOld, ckt->CKTrhsOld, size * sizeof(double), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTrhsOld, size, double, status) CUDAMEMCPYCHECK (ckt->d_CKTrhsOld, size, double, status)
return (OK) ; return (OK) ;
} }
@ -78,9 +78,9 @@ CKTcircuit *ckt
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)(ckt->d_MatrixSize + 1) ; size = (long unsigned int)(ckt->d_MatrixSize + 1) ;
status = cudaMemcpy (ckt->CKTrhsOld, ckt->d_CKTrhsOld, size * sizeof(double), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (ckt->CKTrhsOld, ckt->d_CKTrhsOld, size * sizeof(double), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (ckt->CKTrhsOld, size, double, status) CUDAMEMCPYCHECK (ckt->CKTrhsOld, size, double, status)
return (OK) ; return (OK) ;
} }

112
src/spicelib/analysis/CUSPICE/cucktsetup.c
View File

@ -76,88 +76,88 @@ CKTcircuit *ckt
size2 = (long unsigned int)ckt->CKTnumStates ; size2 = (long unsigned int)ckt->CKTnumStates ;
size3 = (long unsigned int)ckt->total_n_timeSteps ; size3 = (long unsigned int)ckt->total_n_timeSteps ;
/* Topology Matrix Handling */ /* Topology Matrix Handling */
status = cudaMalloc ((void **)&(ckt->CKTmatrix->d_CKTrhs), (n + 1) * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->CKTmatrix->d_CKTrhs), (n + 1) * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->CKTmatrix->d_CKTrhs, (n + 1), double, status) CUDAMALLOCCHECK (ckt->CKTmatrix->d_CKTrhs, (n + 1), double, status)
status = cudaMalloc ((void **)&(ckt->CKTmatrix->d_CKTkluAx), nz * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->CKTmatrix->d_CKTkluAx), nz * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->CKTmatrix->d_CKTkluAx, nz, double, status) CUDAMALLOCCHECK (ckt->CKTmatrix->d_CKTkluAx, nz, double, status)
status = cudaMalloc ((void **)&(ckt->d_CKTloadOutput), m * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTloadOutput), m * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTloadOutput, m, double, status) CUDAMALLOCCHECK (ckt->d_CKTloadOutput, m, double, status)
status = cudaMalloc ((void **)&(ckt->d_CKTloadOutputRHS), mRHS * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTloadOutputRHS), mRHS * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTloadOutputRHS, mRHS, double, status) CUDAMALLOCCHECK (ckt->d_CKTloadOutputRHS, mRHS, double, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRp), (nz + 1) * sizeof(int)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRp), (nz + 1) * sizeof(int)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRp, (nz + 1), int, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRp, (nz + 1), int, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRj), TopologyNNZ * sizeof(int)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRj), TopologyNNZ * sizeof(int)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRj, TopologyNNZ, int, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRj, TopologyNNZ, int, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRx), TopologyNNZ * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRx), TopologyNNZ * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRx, TopologyNNZ, double, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRx, TopologyNNZ, double, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRpRHS), ((n + 1) + 1) * sizeof(int)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRpRHS), ((n + 1) + 1) * sizeof(int)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRpRHS, ((n + 1) + 1), int, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRpRHS, ((n + 1) + 1), int, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRjRHS), TopologyNNZRHS * sizeof(int)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRjRHS), TopologyNNZRHS * sizeof(int)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRjRHS, TopologyNNZRHS, int, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRjRHS, TopologyNNZRHS, int, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRxRHS), TopologyNNZRHS * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtopologyMatrixCSRxRHS), TopologyNNZRHS * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRxRHS, TopologyNNZRHS, double, status) CUDAMALLOCCHECK (ckt->d_CKTtopologyMatrixCSRxRHS, TopologyNNZRHS, double, status)
cudaMemset (ckt->d_CKTloadOutput + ckt->total_n_values, 0, sizeof(double)) ; //DiagGmin is 0 at the beginning cudaMemset (ckt->d_CKTloadOutput + ckt->total_n_values, 0, sizeof(double)) ; //DiagGmin is 0 at the beginning
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRp, ckt->CKTtopologyMatrixCSRp, (nz + 1) * sizeof(int), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRp, ckt->CKTtopologyMatrixCSRp, (nz + 1) * sizeof(int), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRp, (nz + 1), int, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRp, (nz + 1), int, status)
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRj, ckt->CKTtopologyMatrixCOOj, TopologyNNZ * sizeof(int), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRj, ckt->CKTtopologyMatrixCOOj, TopologyNNZ * sizeof(int), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRj, TopologyNNZ, int, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRj, TopologyNNZ, int, status)
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRx, ckt->CKTtopologyMatrixCOOx, TopologyNNZ * sizeof(double), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRx, ckt->CKTtopologyMatrixCOOx, TopologyNNZ * sizeof(double), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRx, TopologyNNZ, double, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRx, TopologyNNZ, double, status)
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRpRHS, ckt->CKTtopologyMatrixCSRpRHS, ((n + 1) + 1) * sizeof(int), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRpRHS, ckt->CKTtopologyMatrixCSRpRHS, ((n + 1) + 1) * sizeof(int), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRpRHS, ((n + 1) + 1), int, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRpRHS, ((n + 1) + 1), int, status)
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRjRHS, ckt->CKTtopologyMatrixCOOjRHS, TopologyNNZRHS * sizeof(int), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRjRHS, ckt->CKTtopologyMatrixCOOjRHS, TopologyNNZRHS * sizeof(int), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRjRHS, TopologyNNZRHS, int, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRjRHS, TopologyNNZRHS, int, status)
status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRxRHS, ckt->CKTtopologyMatrixCOOxRHS, TopologyNNZRHS * sizeof(double), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTtopologyMatrixCSRxRHS, ckt->CKTtopologyMatrixCOOxRHS, TopologyNNZRHS * sizeof(double), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRxRHS, TopologyNNZRHS, double, status) CUDAMEMCPYCHECK (ckt->d_CKTtopologyMatrixCSRxRHS, TopologyNNZRHS, double, status)
/* ------------------------ */ /* ------------------------ */
status = cudaMalloc ((void **)&(ckt->d_CKTnoncon), sizeof(int)) ; status = cudaMalloc ((void **)&(ckt->d_CKTnoncon), sizeof(int)) ;
CUDAMALLOCCHECK (ckt->d_CKTnoncon, 1, int, status) CUDAMALLOCCHECK (ckt->d_CKTnoncon, 1, int, status)
status = cudaMalloc ((void **)&(ckt->d_CKTrhsOld), size1 * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTrhsOld), size1 * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTrhsOld, size1, double, status) CUDAMALLOCCHECK (ckt->d_CKTrhsOld, size1, double, status)
for (i = 0 ; i <= MAX (2, ckt->CKTmaxOrder) + 1 ; i++) /* dctran needs 3 states at least */ for (i = 0 ; i <= MAX (2, ckt->CKTmaxOrder) + 1 ; i++) /* dctran needs 3 states at least */
{ {
status = cudaMalloc ((void **)&(ckt->d_CKTstates[i]), size2 * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTstates[i]), size2 * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTstates[i], size2, double, status) CUDAMALLOCCHECK (ckt->d_CKTstates[i], size2, double, status)
} }
/* Truncation Error */ /* Truncation Error */
status = cudaMalloc ((void **)&(ckt->dD_CKTstates), 8 * sizeof(double *)) ; status = cudaMalloc ((void **)&(ckt->dD_CKTstates), 8 * sizeof(double *)) ;
CUDAMALLOCCHECK (ckt->dD_CKTstates, 8, double *, status) CUDAMALLOCCHECK (ckt->dD_CKTstates, 8, double *, status)
status = cudaMemcpy (ckt->dD_CKTstates, ckt->d_CKTstates, 8 * sizeof(double *), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->dD_CKTstates, ckt->d_CKTstates, 8 * sizeof(double *), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->dD_CKTstates, 8, double *, status) CUDAMEMCPYCHECK (ckt->dD_CKTstates, 8, double *, status)
status = cudaMalloc ((void **)&(ckt->d_CKTdeltaOld), 7 * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTdeltaOld), 7 * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTdeltaOld, 7, double, status) CUDAMALLOCCHECK (ckt->d_CKTdeltaOld, 7, double, status)
// ckt->CKTtimeSteps = (double *) malloc (size3 * sizeof(double)) ; // ckt->CKTtimeSteps = (double *) malloc (size3 * sizeof(double)) ;
status = cudaMalloc ((void **)&(ckt->d_CKTtimeSteps), size3 * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtimeSteps), size3 * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTtimeSteps, size3, double, status) CUDAMALLOCCHECK (ckt->d_CKTtimeSteps, size3, double, status)
status = cudaMalloc ((void **)&(ckt->d_CKTtimeStepsOut), size3 * sizeof(double)) ; status = cudaMalloc ((void **)&(ckt->d_CKTtimeStepsOut), size3 * sizeof(double)) ;
CUDAMALLOCCHECK (ckt->d_CKTtimeStepsOut, size3, double, status) CUDAMALLOCCHECK (ckt->d_CKTtimeStepsOut, size3, double, status)
return (OK) ; return (OK) ;
} }

52
src/spicelib/analysis/CUSPICE/cucktstatesupdate.c
View File

@ -47,8 +47,8 @@ CKTcircuit *ckt
{ {
long unsigned int size ; long unsigned int size ;
size = (long unsigned int)ckt->CKTnumStates ; size = (long unsigned int)ckt->CKTnumStates ;
cudaMemset (ckt->d_CKTstate0, 0, size * sizeof(double)) ; cudaMemset (ckt->d_CKTstate0, 0, size * sizeof(double)) ;
return (OK) ; return (OK) ;
} }
@ -62,9 +62,9 @@ CKTcircuit *ckt
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)ckt->CKTnumStates ; size = (long unsigned int)ckt->CKTnumStates ;
status = cudaMemcpy (ckt->d_CKTstate0, ckt->CKTstate0, size * sizeof(double), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTstate0, ckt->CKTstate0, size * sizeof(double), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTstate0, size, double, status) CUDAMEMCPYCHECK (ckt->d_CKTstate0, size, double, status)
return (OK) ; return (OK) ;
} }
@ -78,9 +78,9 @@ CKTcircuit *ckt
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)ckt->CKTnumStates ; size = (long unsigned int)ckt->CKTnumStates ;
status = cudaMemcpy (ckt->CKTstate0, ckt->d_CKTstate0, size * sizeof(double), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (ckt->CKTstate0, ckt->d_CKTstate0, size * sizeof(double), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (ckt->CKTstate0, size, double, status) CUDAMEMCPYCHECK (ckt->CKTstate0, size, double, status)
return (OK) ; return (OK) ;
} }
@ -91,12 +91,12 @@ cuCKTstate01copy
CKTcircuit *ckt CKTcircuit *ckt
) )
{ {
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)ckt->CKTnumStates ; size = (long unsigned int)ckt->CKTnumStates ;
status = cudaMemcpy (ckt->d_CKTstate1, ckt->d_CKTstate0, size * sizeof(double), cudaMemcpyDeviceToDevice) ; status = cudaMemcpy (ckt->d_CKTstate1, ckt->d_CKTstate0, size * sizeof(double), cudaMemcpyDeviceToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTstate1, size, double, status) CUDAMEMCPYCHECK (ckt->d_CKTstate1, size, double, status)
return (OK) ; return (OK) ;
} }
@ -110,11 +110,11 @@ CKTcircuit *ckt
int i ; int i ;
double *temp ; double *temp ;
temp = ckt->d_CKTstates [ckt->CKTmaxOrder + 1] ; temp = ckt->d_CKTstates [ckt->CKTmaxOrder + 1] ;
for (i = ckt->CKTmaxOrder ; i >= 0 ; i--) for (i = ckt->CKTmaxOrder ; i >= 0 ; i--)
ckt->d_CKTstates [i + 1] = ckt->d_CKTstates [i] ; ckt->d_CKTstates [i + 1] = ckt->d_CKTstates [i] ;
ckt->d_CKTstates [0] = temp ; ckt->d_CKTstates [0] = temp ;
return (OK) ; return (OK) ;
} }
@ -125,16 +125,16 @@ cuCKTstate123copy
CKTcircuit *ckt CKTcircuit *ckt
) )
{ {
long unsigned int size ; long unsigned int size ;
cudaError_t status ; cudaError_t status ;
size = (long unsigned int)ckt->CKTnumStates ; size = (long unsigned int)ckt->CKTnumStates ;
status = cudaMemcpy (ckt->d_CKTstate2, ckt->d_CKTstate1, size * sizeof(double), cudaMemcpyDeviceToDevice) ; status = cudaMemcpy (ckt->d_CKTstate2, ckt->d_CKTstate1, size * sizeof(double), cudaMemcpyDeviceToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTstate2, size, double, status) CUDAMEMCPYCHECK (ckt->d_CKTstate2, size, double, status)
status = cudaMemcpy (ckt->d_CKTstate3, ckt->d_CKTstate1, size * sizeof(double), cudaMemcpyDeviceToDevice) ; status = cudaMemcpy (ckt->d_CKTstate3, ckt->d_CKTstate1, size * sizeof(double), cudaMemcpyDeviceToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTstate3, size, double, status) CUDAMEMCPYCHECK (ckt->d_CKTstate3, size, double, status)
return (OK) ; return (OK) ;
} }
@ -147,8 +147,8 @@ CKTcircuit *ckt
{ {
cudaError_t status ; cudaError_t status ;
status = cudaMemcpy (ckt->d_CKTdeltaOld, ckt->CKTdeltaOld, 7 * sizeof(double), cudaMemcpyHostToDevice) ; status = cudaMemcpy (ckt->d_CKTdeltaOld, ckt->CKTdeltaOld, 7 * sizeof(double), cudaMemcpyHostToDevice) ;
CUDAMEMCPYCHECK (ckt->d_CKTdeltaOld, 7, double, status) CUDAMEMCPYCHECK (ckt->d_CKTdeltaOld, 7, double, status)
return (OK) ; return (OK) ;
} }

16
src/spicelib/analysis/CUSPICE/cucktsystem.c
View File

@ -48,16 +48,16 @@ CKTcircuit *ckt
long unsigned int nz, n ; long unsigned int nz, n ;
cudaError_t status ; cudaError_t status ;
nz = (long unsigned int)ckt->CKTmatrix->CKTklunz ; nz = (long unsigned int)ckt->CKTmatrix->CKTklunz ;
n = (long unsigned int)ckt->CKTmatrix->CKTkluN ; n = (long unsigned int)ckt->CKTmatrix->CKTkluN ;
/* Copy back the Matrix */ /* Copy back the Matrix */
status = cudaMemcpy (ckt->CKTmatrix->CKTkluAx, ckt->CKTmatrix->d_CKTkluAx, nz * sizeof(double), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (ckt->CKTmatrix->CKTkluAx, ckt->CKTmatrix->d_CKTkluAx, nz * sizeof(double), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (ckt->CKTmatrix->CKTkluAx, nz, double, status) CUDAMEMCPYCHECK (ckt->CKTmatrix->CKTkluAx, nz, double, status)
/* Copy back the RHS */ /* Copy back the RHS */
status = cudaMemcpy (ckt->CKTrhs, ckt->CKTmatrix->d_CKTrhs, (n + 1) * sizeof(double), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (ckt->CKTrhs, ckt->CKTmatrix->d_CKTrhs, (n + 1) * sizeof(double), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (ckt->CKTrhs, (n + 1), double, status) CUDAMEMCPYCHECK (ckt->CKTrhs, (n + 1), double, status)
return (OK) ; return (OK) ;
} }

90
src/spicelib/analysis/CUSPICE/cuckttrunc.cu
View File

@ -31,63 +31,63 @@ cuCKTtrunc
CKTcircuit *ckt, double timetemp, double *timeStep CKTcircuit *ckt, double timetemp, double *timeStep
) )
{ {
long unsigned int size ; long unsigned int size ;
double timetempGPU ; double timetempGPU ;
int thread_x, thread_y, block_x ; int thread_x, thread_y, block_x ;
cudaError_t status ; cudaError_t status ;
/* Determining how many blocks should exist in the kernel */ /* Determining how many blocks should exist in the kernel */
thread_x = 1 ; thread_x = 1 ;
thread_y = 256 ; thread_y = 256 ;
if (ckt->total_n_timeSteps % thread_y != 0) if (ckt->total_n_timeSteps % thread_y != 0)
block_x = (int)((ckt->total_n_timeSteps + thread_y - 1) / thread_y) ; block_x = (int)((ckt->total_n_timeSteps + thread_y - 1) / thread_y) ;
else else
block_x = ckt->total_n_timeSteps / thread_y ; block_x = ckt->total_n_timeSteps / thread_y ;
dim3 thread (thread_x, thread_y) ; dim3 thread (thread_x, thread_y) ;
/* Kernel launch */ /* Kernel launch */
status = cudaGetLastError () ; // clear error status status = cudaGetLastError () ; // clear error status
cuCKTtrunc_kernel <<< block_x, thread, thread_y * sizeof(double) >>> (ckt->d_CKTtimeSteps, ckt->d_CKTtimeStepsOut, ckt->total_n_timeSteps) ; cuCKTtrunc_kernel <<< block_x, thread, thread_y * sizeof(double) >>> (ckt->d_CKTtimeSteps, ckt->d_CKTtimeStepsOut, ckt->total_n_timeSteps) ;
cudaDeviceSynchronize () ; cudaDeviceSynchronize () ;
status = cudaGetLastError () ; // check for launch error status = cudaGetLastError () ; // check for launch error
if (status != cudaSuccess) if (status != cudaSuccess)
{ {
fprintf (stderr, "Kernel 1 launch failure in cuCKTtrunc\n\n") ; fprintf (stderr, "Kernel 1 launch failure in cuCKTtrunc\n\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
cuCKTtrunc_kernel <<< 1, thread, thread_y * sizeof(double) >>> (ckt->d_CKTtimeStepsOut, ckt->d_CKTtimeSteps, block_x) ; cuCKTtrunc_kernel <<< 1, thread, thread_y * sizeof(double) >>> (ckt->d_CKTtimeStepsOut, ckt->d_CKTtimeSteps, block_x) ;
cudaDeviceSynchronize () ; cudaDeviceSynchronize () ;
status = cudaGetLastError () ; // check for launch error status = cudaGetLastError () ; // check for launch error
if (status != cudaSuccess) if (status != cudaSuccess)
{ {
fprintf (stderr, "Kernel 2 launch failure in cuCKTtrunc\n\n") ; fprintf (stderr, "Kernel 2 launch failure in cuCKTtrunc\n\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
/* Copy back the reduction result */ /* Copy back the reduction result */
size = (long unsigned int)(1) ; size = (long unsigned int)(1) ;
status = cudaMemcpy (&timetempGPU, ckt->d_CKTtimeSteps, size * sizeof(double), cudaMemcpyDeviceToHost) ; status = cudaMemcpy (&timetempGPU, ckt->d_CKTtimeSteps, size * sizeof(double), cudaMemcpyDeviceToHost) ;
CUDAMEMCPYCHECK (&timetempGPU, size, double, status) CUDAMEMCPYCHECK (&timetempGPU, size, double, status)
/* Final Comparison */ /* Final Comparison */
if (timetempGPU < timetemp) if (timetempGPU < timetemp)
{ {
timetemp = timetempGPU ; timetemp = timetempGPU ;
} }
if (2 * *timeStep < timetemp) if (2 * *timeStep < timetemp)
{ {
*timeStep = 2 * *timeStep ; *timeStep = 2 * *timeStep ;
} else { } else {
*timeStep = timetemp ; *timeStep = timetemp ;
} }
return 0 ; return 0 ;
} }

76
src/spicelib/analysis/cktload.c
View File

@ -127,50 +127,50 @@ CKTload(CKTcircuit *ckt)
} }
#ifdef USE_CUSPICE #ifdef USE_CUSPICE
/* Copy the CKTdiagGmin value to the GPU */ /* Copy the CKTdiagGmin value to the GPU */
// The real Gmin is needed only when the matrix will reside entirely on the GPU // The real Gmin is needed only when the matrix will reside entirely on the GPU
// Right now, only some models support CUDA, so the matrix is only partially created on the GPU // Right now, only some models support CUDA, so the matrix is only partially created on the GPU
cudaMemset (ckt->d_CKTloadOutput + ckt->total_n_values, 0, sizeof(double)) ; cudaMemset (ckt->d_CKTloadOutput + ckt->total_n_values, 0, sizeof(double)) ;
//cudaError_t statusMemcpy ; //cudaError_t statusMemcpy ;
//statusMemcpy = cudaMemcpy (ckt->d_CKTloadOutput + ckt->total_n_values, &(ckt->CKTdiagGmin), sizeof(double), cudaMemcpyHostToDevice) ; //statusMemcpy = cudaMemcpy (ckt->d_CKTloadOutput + ckt->total_n_values, &(ckt->CKTdiagGmin), sizeof(double), cudaMemcpyHostToDevice) ;
//CUDAMEMCPYCHECK (ckt->d_CKTloadOutput + ckt->total_n_values, 1, double, statusMemcpy) //CUDAMEMCPYCHECK (ckt->d_CKTloadOutput + ckt->total_n_values, 1, double, statusMemcpy)
/* Performing CSRMV for the Sparse Matrix using CUSPARSE */ /* Performing CSRMV for the Sparse Matrix using CUSPARSE */
cusparseStatus = cusparseDcsrmv ((cusparseHandle_t)(ckt->CKTmatrix->CKTcsrmvHandle), cusparseStatus = cusparseDcsrmv ((cusparseHandle_t)(ckt->CKTmatrix->CKTcsrmvHandle),
CUSPARSE_OPERATION_NON_TRANSPOSE, CUSPARSE_OPERATION_NON_TRANSPOSE,
ckt->CKTmatrix->CKTklunz, ckt->total_n_values + 1, ckt->CKTmatrix->CKTklunz, ckt->total_n_values + 1,
ckt->total_n_Ptr + ckt->CKTdiagElements, ckt->total_n_Ptr + ckt->CKTdiagElements,
&alpha, (cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), &alpha, (cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr),
ckt->d_CKTtopologyMatrixCSRx, ckt->d_CKTtopologyMatrixCSRp, ckt->d_CKTtopologyMatrixCSRx, ckt->d_CKTtopologyMatrixCSRp,
ckt->d_CKTtopologyMatrixCSRj, ckt->d_CKTloadOutput, &beta, ckt->d_CKTtopologyMatrixCSRj, ckt->d_CKTloadOutput, &beta,
ckt->CKTmatrix->d_CKTkluAx) ; ckt->CKTmatrix->d_CKTkluAx) ;
if (cusparseStatus != CUSPARSE_STATUS_SUCCESS) if (cusparseStatus != CUSPARSE_STATUS_SUCCESS)
{ {
fprintf (stderr, "CUSPARSE MATRIX Call Error\n") ; fprintf (stderr, "CUSPARSE MATRIX Call Error\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
/* Performing CSRMV for the RHS using CUSPARSE */ /* Performing CSRMV for the RHS using CUSPARSE */
cusparseStatus = cusparseDcsrmv ((cusparseHandle_t)(ckt->CKTmatrix->CKTcsrmvHandle), cusparseStatus = cusparseDcsrmv ((cusparseHandle_t)(ckt->CKTmatrix->CKTcsrmvHandle),
CUSPARSE_OPERATION_NON_TRANSPOSE, CUSPARSE_OPERATION_NON_TRANSPOSE,
ckt->CKTmatrix->CKTkluN + 1, ckt->total_n_valuesRHS, ckt->total_n_PtrRHS, ckt->CKTmatrix->CKTkluN + 1, ckt->total_n_valuesRHS, ckt->total_n_PtrRHS,
&alpha, (cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), &alpha, (cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr),
ckt->d_CKTtopologyMatrixCSRxRHS, ckt->d_CKTtopologyMatrixCSRpRHS, ckt->d_CKTtopologyMatrixCSRxRHS, ckt->d_CKTtopologyMatrixCSRpRHS,
ckt->d_CKTtopologyMatrixCSRjRHS, ckt->d_CKTloadOutputRHS, &beta, ckt->d_CKTtopologyMatrixCSRjRHS, ckt->d_CKTloadOutputRHS, &beta,
ckt->CKTmatrix->d_CKTrhs) ; ckt->CKTmatrix->d_CKTrhs) ;
if (cusparseStatus != CUSPARSE_STATUS_SUCCESS) if (cusparseStatus != CUSPARSE_STATUS_SUCCESS)
{ {
fprintf (stderr, "CUSPARSE RHS Call Error\n") ; fprintf (stderr, "CUSPARSE RHS Call Error\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
cudaDeviceSynchronize () ; cudaDeviceSynchronize () ;
status = cuCKTsystemDtoH (ckt) ; status = cuCKTsystemDtoH (ckt) ;
if (status != 0) if (status != 0)
return (E_NOMEM) ; return (E_NOMEM) ;
#endif #endif
#ifdef XSPICE #ifdef XSPICE

42
src/spicelib/analysis/cktsetup.c
View File

@ -337,30 +337,30 @@ CKTsetup(CKTcircuit *ckt)
} }
#ifdef USE_CUSPICE #ifdef USE_CUSPICE
ckt->d_MatrixSize = SMPmatSize (ckt->CKTmatrix) ; ckt->d_MatrixSize = SMPmatSize (ckt->CKTmatrix) ;
status = cuCKTsetup (ckt) ; status = cuCKTsetup (ckt) ;
if (status != 0) if (status != 0)
return (E_NOMEM) ; return (E_NOMEM) ;
/* CUSPARSE Handle Creation */ /* CUSPARSE Handle Creation */
cusparseStatus = cusparseCreate ((cusparseHandle_t *)(&(ckt->CKTmatrix->CKTcsrmvHandle))) ; cusparseStatus = cusparseCreate ((cusparseHandle_t *)(&(ckt->CKTmatrix->CKTcsrmvHandle))) ;
if (cusparseStatus != CUSPARSE_STATUS_SUCCESS) if (cusparseStatus != CUSPARSE_STATUS_SUCCESS)
{ {
fprintf (stderr, "CUSPARSE Handle Setup Error\n") ; fprintf (stderr, "CUSPARSE Handle Setup Error\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
/* CUSPARSE Matrix Descriptor Creation */ /* CUSPARSE Matrix Descriptor Creation */
cusparseStatus = cusparseCreateMatDescr ((cusparseMatDescr_t *)(&(ckt->CKTmatrix->CKTcsrmvDescr))) ; cusparseStatus = cusparseCreateMatDescr ((cusparseMatDescr_t *)(&(ckt->CKTmatrix->CKTcsrmvDescr))) ;
if (cusparseStatus != CUSPARSE_STATUS_SUCCESS) if (cusparseStatus != CUSPARSE_STATUS_SUCCESS)
{ {
fprintf (stderr, "CUSPARSE Matrix Descriptor Setup Error\n") ; fprintf (stderr, "CUSPARSE Matrix Descriptor Setup Error\n") ;
return (E_NOMEM) ; return (E_NOMEM) ;
} }
/* CUSPARSE Matrix Properties Definition */ /* CUSPARSE Matrix Properties Definition */
cusparseSetMatType ((cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), CUSPARSE_MATRIX_TYPE_GENERAL) ; cusparseSetMatType ((cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), CUSPARSE_MATRIX_TYPE_GENERAL) ;
cusparseSetMatIndexBase ((cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), CUSPARSE_INDEX_BASE_ZERO) ; cusparseSetMatIndexBase ((cusparseMatDescr_t)(ckt->CKTmatrix->CKTcsrmvDescr), CUSPARSE_INDEX_BASE_ZERO) ;
#endif #endif
#ifdef WANT_SENSE2 #ifdef WANT_SENSE2