Multiply( const matrix_type & A ,
const size_type nrow ,
const size_type ncol ,
const vector_type & x ,
const vector_type & y )
{
CudaSparseSingleton & s = CudaSparseSingleton::singleton();
const scalar_type alpha = 1 , beta = 0 ;
cusparseStatus_t status =
cusparseDcsrmv( s.handle ,
CUSPARSE_OPERATION_NON_TRANSPOSE ,
nrow , ncol , A.coefficients.dimension_0() ,
&alpha ,
s.descra ,
A.coefficients.ptr_on_device() ,
A.graph.row_map.ptr_on_device() ,
A.graph.entries.ptr_on_device() ,
x.ptr_on_device() ,
&beta ,
y.ptr_on_device() );
if ( CUSPARSE_STATUS_SUCCESS != status ) {
throw std::runtime_error( std::string("ERROR - cusparseDcsrmv " ) );
}
}
inline cusparseStatus_t cusparseTcsrmv (
cusparseHandle_t handle,
cusparseOperation_t transA,
int m,
int n,
int nnz,
const double *alpha,
const cusparseMatDescr_t descrA,
const double *csrValA,
const int *csrRowPtrA,
const int *csrColIndA,
const double *x,
const double *beta,
double *y
)
{
return cusparseDcsrmv (
handle,
transA,
m,
n,
nnz,
alpha,
descrA,
csrValA,
csrRowPtrA,
csrColIndA,
x,
beta,
y
);
}
static void apply( const matrix_type & A ,
const vector_type & x ,
const vector_type & y )
{
CudaSparseSingleton & s = CudaSparseSingleton::singleton();
const double alpha = 1 , beta = 0 ;
const int n = A.graph.row_map.dimension_0() - 1 ;
const int nz = A.graph.entries.dimension_0();
cusparseStatus_t status =
cusparseDcsrmv( s.handle ,
CUSPARSE_OPERATION_NON_TRANSPOSE ,
n , n , nz ,
&alpha ,
s.descra ,
A.values.ptr_on_device() ,
A.graph.row_map.ptr_on_device() ,
A.graph.entries.ptr_on_device() ,
x.ptr_on_device() ,
&beta ,
y.ptr_on_device() );
if ( CUSPARSE_STATUS_SUCCESS != status ) {
throw std::runtime_error( std::string("ERROR - cusparseDcsrmv " ) );
}
}
void mul(const device_vector<double> &x, device_vector<double> &y,
double alpha = 1, bool append = false) const
{
double beta = append ? 1.0 : 0.0;
cuda_check(
cusparseDcsrmv(handle, CUSPARSE_OPERATION_NON_TRANSPOSE,
n, m, nnz, &alpha, desc.get(),
val.raw_ptr(), row.raw_ptr(), col.raw_ptr(),
x.raw_ptr(), &beta, y.raw_ptr()
)
);
}
void cusparse_wrapper::matvec(const sparse_matrix* A, const double* x, double* y,
bool transpose, double alpha, double beta)
{
// Flip the tranpose option, since we actually have the tranpose stored
cusparseOperation_t transA = transpose ? CUSPARSE_OPERATION_NON_TRANSPOSE :
CUSPARSE_OPERATION_TRANSPOSE;
cusparseStatus_t status =
cusparseDcsrmv(handle,
transA, A->m, A->n, A->nnz,
&alpha,
A->descrA, A->csrValA, A->csrRowPtrA, A->csrColIndA,
x,
&beta,
y
);
assert(status == CUSPARSE_STATUS_SUCCESS);
}
extern "C" magma_int_t
magma_d_spmv(
double alpha,
magma_d_matrix A,
magma_d_matrix x,
double beta,
magma_d_matrix y,
magma_queue_t queue )
{
magma_int_t info = 0;
magma_d_matrix x2={Magma_CSR};
cusparseHandle_t cusparseHandle = 0;
cusparseMatDescr_t descr = 0;
// make sure RHS is a dense matrix
if ( x.storage_type != Magma_DENSE ) {
printf("error: only dense vectors are supported for SpMV.\n");
info = MAGMA_ERR_NOT_SUPPORTED;
goto cleanup;
}
if ( A.memory_location != x.memory_location ||
x.memory_location != y.memory_location ) {
printf("error: linear algebra objects are not located in same memory!\n");
printf("memory locations are: %d %d %d\n",
A.memory_location, x.memory_location, y.memory_location );
info = MAGMA_ERR_INVALID_PTR;
goto cleanup;
}
// DEV case
if ( A.memory_location == Magma_DEV ) {
if ( A.num_cols == x.num_rows && x.num_cols == 1 ) {
if ( A.storage_type == Magma_CSR || A.storage_type == Magma_CUCSR
|| A.storage_type == Magma_CSRL
|| A.storage_type == Magma_CSRU ) {
CHECK_CUSPARSE( cusparseCreate( &cusparseHandle ));
CHECK_CUSPARSE( cusparseSetStream( cusparseHandle, queue->cuda_stream() ));
CHECK_CUSPARSE( cusparseCreateMatDescr( &descr ));
CHECK_CUSPARSE( cusparseSetMatType( descr, CUSPARSE_MATRIX_TYPE_GENERAL ));
CHECK_CUSPARSE( cusparseSetMatIndexBase( descr, CUSPARSE_INDEX_BASE_ZERO ));
cusparseDcsrmv( cusparseHandle,CUSPARSE_OPERATION_NON_TRANSPOSE,
A.num_rows, A.num_cols, A.nnz, &alpha, descr,
A.dval, A.drow, A.dcol, x.dval, &beta, y.dval );
}
else if ( A.storage_type == Magma_ELL ) {
//printf("using ELLPACKT kernel for SpMV: ");
CHECK( magma_dgeelltmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.dval, A.dcol, x.dval, beta,
y.dval, queue ));
//printf("done.\n");
}
else if ( A.storage_type == Magma_ELLPACKT ) {
//printf("using ELL kernel for SpMV: ");
CHECK( magma_dgeellmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.dval, A.dcol, x.dval, beta,
y.dval, queue ));
//printf("done.\n");
}
else if ( A.storage_type == Magma_ELLRT ) {
//printf("using ELLRT kernel for SpMV: ");
CHECK( magma_dgeellrtmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.dval, A.dcol, A.drow, x.dval,
beta, y.dval, A.alignment, A.blocksize, queue ));
//printf("done.\n");
}
else if ( A.storage_type == Magma_SELLP ) {
//printf("using SELLP kernel for SpMV: ");
CHECK( magma_dgesellpmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.blocksize, A.numblocks, A.alignment,
alpha, A.dval, A.dcol, A.drow, x.dval, beta, y.dval, queue ));
//printf("done.\n");
}
else if ( A.storage_type == Magma_DENSE ) {
//printf("using DENSE kernel for SpMV: ");
magmablas_dgemv( MagmaNoTrans, A.num_rows, A.num_cols, alpha,
A.dval, A.num_rows, x.dval, 1, beta, y.dval,
1, queue );
//printf("done.\n");
}
else if ( A.storage_type == Magma_SPMVFUNCTION ) {
//printf("using DENSE kernel for SpMV: ");
CHECK( magma_dcustomspmv( alpha, x, beta, y, queue ));
//printf("done.\n");
}
else if ( A.storage_type == Magma_BCSR ) {
//printf("using CUSPARSE BCSR kernel for SpMV: ");
// CUSPARSE context //
cusparseDirection_t dirA = CUSPARSE_DIRECTION_ROW;
int mb = magma_ceildiv( A.num_rows, A.blocksize );
int nb = magma_ceildiv( A.num_cols, A.blocksize );
CHECK_CUSPARSE( cusparseCreate( &cusparseHandle ));
CHECK_CUSPARSE( cusparseSetStream( cusparseHandle, queue->cuda_stream() ));
CHECK_CUSPARSE( cusparseCreateMatDescr( &descr ));
cusparseDbsrmv( cusparseHandle, dirA,
CUSPARSE_OPERATION_NON_TRANSPOSE, mb, nb, A.numblocks,
&alpha, descr, A.dval, A.drow, A.dcol, A.blocksize, x.dval,
&beta, y.dval );
}
else {
printf("error: format not supported.\n");
info = MAGMA_ERR_NOT_SUPPORTED;
}
}
else if ( A.num_cols < x.num_rows || x.num_cols > 1 ) {
magma_int_t num_vecs = x.num_rows / A.num_cols * x.num_cols;
if ( A.storage_type == Magma_CSR ) {
CHECK_CUSPARSE( cusparseCreate( &cusparseHandle ));
CHECK_CUSPARSE( cusparseSetStream( cusparseHandle, queue->cuda_stream() ));
CHECK_CUSPARSE( cusparseCreateMatDescr( &descr ));
CHECK_CUSPARSE( cusparseSetMatType( descr, CUSPARSE_MATRIX_TYPE_GENERAL ));
CHECK_CUSPARSE( cusparseSetMatIndexBase( descr, CUSPARSE_INDEX_BASE_ZERO ));
if ( x.major == MagmaColMajor) {
cusparseDcsrmm(cusparseHandle,
CUSPARSE_OPERATION_NON_TRANSPOSE,
A.num_rows, num_vecs, A.num_cols, A.nnz,
&alpha, descr, A.dval, A.drow, A.dcol,
x.dval, A.num_cols, &beta, y.dval, A.num_cols);
} else if ( x.major == MagmaRowMajor) {
/*cusparseDcsrmm2(cusparseHandle,
CUSPARSE_OPERATION_NON_TRANSPOSE,
CUSPARSE_OPERATION_TRANSPOSE,
A.num_rows, num_vecs, A.num_cols, A.nnz,
&alpha, descr, A.dval, A.drow, A.dcol,
x.dval, A.num_cols, &beta, y.dval, A.num_cols);
*/
}
} else if ( A.storage_type == Magma_SELLP ) {
if ( x.major == MagmaRowMajor) {
CHECK( magma_dmgesellpmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, A.blocksize, A.numblocks, A.alignment,
alpha, A.dval, A.dcol, A.drow, x.dval, beta, y.dval, queue ));
}
else if ( x.major == MagmaColMajor) {
// transpose first to row major
CHECK( magma_dvtranspose( x, &x2, queue ));
CHECK( magma_dmgesellpmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, A.blocksize, A.numblocks, A.alignment,
alpha, A.dval, A.dcol, A.drow, x2.dval, beta, y.dval, queue ));
}
}
/*if ( A.storage_type == Magma_DENSE ) {
//printf("using DENSE kernel for SpMV: ");
magmablas_dmgemv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, alpha, A.dval, A.num_rows, x.dval, 1,
beta, y.dval, 1 );
//printf("done.\n");
}*/
else {
printf("error: format not supported.\n");
info = MAGMA_ERR_NOT_SUPPORTED;
}
}
}
// CPU case missing!
else {
printf("error: CPU not yet supported.\n");
info = MAGMA_ERR_NOT_SUPPORTED;
}
cleanup:
cusparseDestroyMatDescr( descr );
cusparseDestroy( cusparseHandle );
cusparseHandle = 0;
descr = 0;
magma_dmfree(&x2, queue );
return info;
}
/* ////////////////////////////////////////////////////////////////////////////
-- testing sparse matrix vector product
*/
int main( int argc, char** argv )
{
TESTING_INIT();
magma_queue_t queue;
magma_queue_create( /*devices[ opts->device ],*/ &queue );
magma_d_sparse_matrix hA, hA_SELLP, hA_ELL, dA, dA_SELLP, dA_ELL;
hA_SELLP.blocksize = 8;
hA_SELLP.alignment = 8;
real_Double_t start, end, res;
magma_int_t *pntre;
double c_one = MAGMA_D_MAKE(1.0, 0.0);
double c_zero = MAGMA_D_MAKE(0.0, 0.0);
magma_int_t i, j;
for( i = 1; i < argc; ++i ) {
if ( strcmp("--blocksize", argv[i]) == 0 ) {
hA_SELLP.blocksize = atoi( argv[++i] );
} else if ( strcmp("--alignment", argv[i]) == 0 ) {
hA_SELLP.alignment = atoi( argv[++i] );
} else
break;
}
printf( "\n# usage: ./run_dspmv"
" [ --blocksize %d --alignment %d (for SELLP) ]"
" matrices \n\n", (int) hA_SELLP.blocksize, (int) hA_SELLP.alignment );
while( i < argc ) {
if ( strcmp("LAPLACE2D", argv[i]) == 0 && i+1 < argc ) { // Laplace test
i++;
magma_int_t laplace_size = atoi( argv[i] );
magma_dm_5stencil( laplace_size, &hA, queue );
} else { // file-matrix test
magma_d_csr_mtx( &hA, argv[i], queue );
}
printf( "\n# matrix info: %d-by-%d with %d nonzeros\n\n",
(int) hA.num_rows,(int) hA.num_cols,(int) hA.nnz );
real_Double_t FLOPS = 2.0*hA.nnz/1e9;
magma_d_vector hx, hy, dx, dy, hrefvec, hcheck;
// init CPU vectors
magma_d_vinit( &hx, Magma_CPU, hA.num_rows, c_zero, queue );
magma_d_vinit( &hy, Magma_CPU, hA.num_rows, c_zero, queue );
// init DEV vectors
magma_d_vinit( &dx, Magma_DEV, hA.num_rows, c_one, queue );
magma_d_vinit( &dy, Magma_DEV, hA.num_rows, c_zero, queue );
#ifdef MAGMA_WITH_MKL
// calling MKL with CSR
pntre = (magma_int_t*)malloc( (hA.num_rows+1)*sizeof(magma_int_t) );
pntre[0] = 0;
for (j=0; j<hA.num_rows; j++ ) {
pntre[j] = hA.row[j+1];
}
MKL_INT num_rows = hA.num_rows;
MKL_INT num_cols = hA.num_cols;
MKL_INT nnz = hA.nnz;
MKL_INT *col;
TESTING_MALLOC_CPU( col, MKL_INT, nnz );
for( magma_int_t t=0; t < hA.nnz; ++t ) {
col[ t ] = hA.col[ t ];
}
MKL_INT *row;
TESTING_MALLOC_CPU( row, MKL_INT, num_rows );
for( magma_int_t t=0; t < hA.num_rows; ++t ) {
row[ t ] = hA.col[ t ];
}
start = magma_wtime();
for (j=0; j<10; j++ ) {
mkl_dcsrmv( "N", &num_rows, &num_cols,
MKL_ADDR(&c_one), "GFNC", MKL_ADDR(hA.val),
col, row, pntre,
MKL_ADDR(hx.val),
MKL_ADDR(&c_zero), MKL_ADDR(hy.val) );
}
end = magma_wtime();
printf( "\n > MKL : %.2e seconds %.2e GFLOP/s (CSR).\n",
(end-start)/10, FLOPS*10/(end-start) );
TESTING_FREE_CPU( row );
TESTING_FREE_CPU( col );
free(pntre);
#endif // MAGMA_WITH_MKL
// copy matrix to GPU
magma_d_mtransfer( hA, &dA, Magma_CPU, Magma_DEV, queue );
// SpMV on GPU (CSR) -- this is the reference!
start = magma_sync_wtime( queue );
for (j=0; j<10; j++)
magma_d_spmv( c_one, dA, dx, c_zero, dy, queue );
end = magma_sync_wtime( queue );
printf( " > MAGMA: %.2e seconds %.2e GFLOP/s (standard CSR).\n",
(end-start)/10, FLOPS*10/(end-start) );
magma_d_mfree(&dA, queue );
magma_d_vtransfer( dy, &hrefvec , Magma_DEV, Magma_CPU, queue );
// convert to ELL and copy to GPU
magma_d_mconvert( hA, &hA_ELL, Magma_CSR, Magma_ELL, queue );
magma_d_mtransfer( hA_ELL, &dA_ELL, Magma_CPU, Magma_DEV, queue );
magma_d_mfree(&hA_ELL, queue );
magma_d_vfree( &dy, queue );
magma_d_vinit( &dy, Magma_DEV, hA.num_rows, c_zero, queue );
// SpMV on GPU (ELL)
start = magma_sync_wtime( queue );
for (j=0; j<10; j++)
magma_d_spmv( c_one, dA_ELL, dx, c_zero, dy, queue );
end = magma_sync_wtime( queue );
printf( " > MAGMA: %.2e seconds %.2e GFLOP/s (standard ELL).\n",
(end-start)/10, FLOPS*10/(end-start) );
magma_d_mfree(&dA_ELL, queue );
magma_d_vtransfer( dy, &hcheck , Magma_DEV, Magma_CPU, queue );
res = 0.0;
for(magma_int_t k=0; k<hA.num_rows; k++ )
res=res + MAGMA_D_REAL(hcheck.val[k]) - MAGMA_D_REAL(hrefvec.val[k]);
if ( res < .000001 )
printf("# tester spmv ELL: ok\n");
else
printf("# tester spmv ELL: failed\n");
magma_d_vfree( &hcheck, queue );
// convert to SELLP and copy to GPU
magma_d_mconvert( hA, &hA_SELLP, Magma_CSR, Magma_SELLP, queue );
magma_d_mtransfer( hA_SELLP, &dA_SELLP, Magma_CPU, Magma_DEV, queue );
magma_d_mfree(&hA_SELLP, queue );
magma_d_vfree( &dy, queue );
magma_d_vinit( &dy, Magma_DEV, hA.num_rows, c_zero, queue );
// SpMV on GPU (SELLP)
start = magma_sync_wtime( queue );
for (j=0; j<10; j++)
magma_d_spmv( c_one, dA_SELLP, dx, c_zero, dy, queue );
end = magma_sync_wtime( queue );
printf( " > MAGMA: %.2e seconds %.2e GFLOP/s (SELLP).\n",
(end-start)/10, FLOPS*10/(end-start) );
magma_d_vtransfer( dy, &hcheck , Magma_DEV, Magma_CPU, queue );
res = 0.0;
for(magma_int_t k=0; k<hA.num_rows; k++ )
res=res + MAGMA_D_REAL(hcheck.val[k]) - MAGMA_D_REAL(hrefvec.val[k]);
printf("# |x-y|_F = %8.2e\n", res);
if ( res < .000001 )
printf("# tester spmv SELL-P: ok\n");
else
printf("# tester spmv SELL-P: failed\n");
magma_d_vfree( &hcheck, queue );
magma_d_mfree(&dA_SELLP, queue );
// SpMV on GPU (CUSPARSE - CSR)
// CUSPARSE context //
cusparseHandle_t cusparseHandle = 0;
cusparseStatus_t cusparseStatus;
cusparseStatus = cusparseCreate(&cusparseHandle);
cusparseSetStream( cusparseHandle, queue );
cusparseMatDescr_t descr = 0;
cusparseStatus = cusparseCreateMatDescr(&descr);
cusparseSetMatType(descr,CUSPARSE_MATRIX_TYPE_GENERAL);
cusparseSetMatIndexBase(descr,CUSPARSE_INDEX_BASE_ZERO);
double alpha = c_one;
double beta = c_zero;
magma_d_vfree( &dy, queue );
magma_d_vinit( &dy, Magma_DEV, hA.num_rows, c_zero, queue );
// copy matrix to GPU
magma_d_mtransfer( hA, &dA, Magma_CPU, Magma_DEV, queue );
start = magma_sync_wtime( queue );
for (j=0; j<10; j++)
cusparseStatus =
cusparseDcsrmv(cusparseHandle,CUSPARSE_OPERATION_NON_TRANSPOSE,
hA.num_rows, hA.num_cols, hA.nnz, &alpha, descr,
dA.dval, dA.drow, dA.dcol, dx.dval, &beta, dy.dval);
end = magma_sync_wtime( queue );
if (cusparseStatus != 0) printf("error in cuSPARSE CSR\n");
printf( " > CUSPARSE: %.2e seconds %.2e GFLOP/s (CSR).\n",
(end-start)/10, FLOPS*10/(end-start) );
cusparseMatDescr_t descrA;
cusparseStatus = cusparseCreateMatDescr(&descrA);
if (cusparseStatus != 0) printf("error\n");
cusparseHybMat_t hybA;
cusparseStatus = cusparseCreateHybMat( &hybA );
if (cusparseStatus != 0) printf("error\n");
magma_d_vtransfer( dy, &hcheck , Magma_DEV, Magma_CPU, queue );
res = 0.0;
for(magma_int_t k=0; k<hA.num_rows; k++ )
res=res + MAGMA_D_REAL(hcheck.val[k]) - MAGMA_D_REAL(hrefvec.val[k]);
printf("# |x-y|_F = %8.2e\n", res);
if ( res < .000001 )
printf("# tester spmv cuSPARSE CSR: ok\n");
else
printf("# tester spmv cuSPARSE CSR: failed\n");
magma_d_vfree( &hcheck, queue );
magma_d_vfree( &dy, queue );
magma_d_vinit( &dy, Magma_DEV, hA.num_rows, c_zero, queue );
cusparseDcsr2hyb(cusparseHandle, hA.num_rows, hA.num_cols,
descrA, dA.dval, dA.drow, dA.dcol,
hybA, 0, CUSPARSE_HYB_PARTITION_AUTO);
start = magma_sync_wtime( queue );
for (j=0; j<10; j++)
cusparseStatus =
cusparseDhybmv( cusparseHandle, CUSPARSE_OPERATION_NON_TRANSPOSE,
&alpha, descrA, hybA,
dx.dval, &beta, dy.dval);
end = magma_sync_wtime( queue );
if (cusparseStatus != 0) printf("error in cuSPARSE HYB\n");
printf( " > CUSPARSE: %.2e seconds %.2e GFLOP/s (HYB).\n",
(end-start)/10, FLOPS*10/(end-start) );
magma_d_vtransfer( dy, &hcheck , Magma_DEV, Magma_CPU, queue );
res = 0.0;
for(magma_int_t k=0; k<hA.num_rows; k++ )
res=res + MAGMA_D_REAL(hcheck.val[k]) - MAGMA_D_REAL(hrefvec.val[k]);
printf("# |x-y|_F = %8.2e\n", res);
if ( res < .000001 )
printf("# tester spmv cuSPARSE HYB: ok\n");
else
printf("# tester spmv cuSPARSE HYB: failed\n");
magma_d_vfree( &hcheck, queue );
cusparseDestroyMatDescr( descrA );
cusparseDestroyHybMat( hybA );
cusparseDestroy( cusparseHandle );
magma_d_mfree(&dA, queue );
printf("\n\n");
// free CPU memory
magma_d_mfree(&hA, queue );
magma_d_vfree(&hx, queue );
magma_d_vfree(&hy, queue );
magma_d_vfree(&hrefvec, queue );
// free GPU memory
magma_d_vfree(&dx, queue );
magma_d_vfree(&dy, queue );
i++;
}
magma_queue_destroy( queue );
TESTING_FINALIZE();
return 0;
}
int main(int argc, char* argv[])
{
const int bufsize = 512;
char buffer[bufsize];
int m,n,S;
double time_st,time_end,time_avg;
//omp_set_num_threads(2);
// printf("\n-----------------\nnumber of threads fired = %d\n-----------------\n",(int)omp_get_num_threads());
if(argc!=2)
{
cout<<"Insufficient arguments"<<endl;
return 1;
}
graph G;
cerr<<"Start reading ";
// time_st=dsecnd();
G.create_graph(argv[1]);
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cout<<"Success "<<endl;
// cerr<<"Reading time "<<time_avg<<endl;
cerr<<"Constructing Matrices ";
// time_st=dsecnd();
G.construct_MNA();
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cerr<<"Done "<<time_avg<<endl;
// G.construct_sparse_MNA();
m=G.node_array.size()-1;
n=G.voltage_edge_id.size();
cout<<endl;
cout<<"MATRIX STAT:"<<endl;
cout<<"Nonzero elements: "<<G.nonzero<<endl;
cout<<"Number of Rows: "<<m+n<<endl;
printf("\n Nonzero = %ld", G.nonzero);
printf("\n Rows = %d", m+n);
cout<<"MAT val: "<<endl;
int i,j;
// G.Mat_val[0] +=100;
/*
for(i=0;i<G.nonzero;i++)
cout<<" "<<G.Mat_val[i];
cout<<endl;
for(i=0;i<G.nonzero;i++)
cout<<" "<<G.columns[i];
cout<<endl;
for(i=0;i<m+n+1;i++)
cout<<" "<<G.rowIndex[i];
cout<<endl;
for(i=0;i<m+n;i++)
{
cout<<endl;
int startindex=G.rowIndex[i];
int endindex=G.rowIndex[i+1];
for(j=startindex;j<endindex;j++)
cout<<" "<<G.Mat_val[j];
cout<<endl;
}
*/
/* for (i=0;i<m+n+1;i++)
{
//cout<<endl;
if(G.rowIndex[i]==G.rowIndex[i+1])
break;
for(j=G.rowIndex[i];j<G.rowIndex[i+1];j++)
{
if(G.Mat_val[j]>10)
cout<<G.Mat_val[j]<<"\t";
}
//cout<<endl;
/*for(j=G.rowIndex[i];j<G.rowIndex[i+1];j++)
{
cout<<G.columns[j]<<"\t";
}
//cout<<endl;
}
cout<<endl;
*/
//printing the matrix
printf("\n Fine till here");
printf("\n");
// int* rowmIndex=(int*)calloc(m+1,sizeof(int));
printf("\n Fine till here");
printf("\n");
//int rowmIndex[5]={1,2,3,4,5};
/* for(i=0;i<m+1;i++)
{
rowmIndex[i]=G.rowIndex[i];
printf(" %d", rowmIndex[i]);
}
*/
cerr<<"Solving Equations "<<endl;
double r1, b, alpha, alpham1, beta, r0, a, na;
const double tol = 0.1;
const int max_iter = 1000000;
int *d_col, *d_row;
double *d_val, *d_x, dot;
double *d_r, *d_p, *d_Ax;
int k;
cublasHandle_t cublasHandle = 0;
cublasStatus_t cublasStatus;
cublasStatus = cublasCreate(&cublasHandle);
checkCudaErrors(cublasStatus);
/* Get handle to the CUSPARSE context */
cusparseHandle_t cusparseHandle = 0;
cusparseStatus_t cusparseStatus;
cusparseStatus = cusparseCreate(&cusparseHandle);
checkCudaErrors(cusparseStatus);
cusparseMatDescr_t descr = 0;
cusparseStatus = cusparseCreateMatDescr(&descr);
checkCudaErrors(cusparseStatus);
cusparseSetMatType(descr,CUSPARSE_MATRIX_TYPE_GENERAL);
cusparseSetMatIndexBase(descr,CUSPARSE_INDEX_BASE_ZERO);
checkCudaErrors(cudaMalloc((void **)&d_col, G.nonzero*sizeof(int)));
checkCudaErrors(cudaMalloc((void **)&d_row, (m+n+1)*sizeof(int)));
checkCudaErrors(cudaMalloc((void **)&d_val, G.nonzero*sizeof(double)));
checkCudaErrors(cudaMalloc((void **)&d_x, (m+n)*sizeof(double)));
checkCudaErrors(cudaMalloc((void **)&d_r, (m+n)*sizeof(double)));
checkCudaErrors(cudaMalloc((void **)&d_p, (m+n)*sizeof(double)));
checkCudaErrors(cudaMalloc((void **)&d_Ax, (m+n)*sizeof(double)));
cudaMemcpy(d_col, G.columns, G.nonzero*sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_row, G.rowIndex, (m+n+1)*sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(d_val, G.Mat_val, G.nonzero*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(d_x, G.x, (m+n)*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(d_r, G.b, (m+n)*sizeof(double), cudaMemcpyHostToDevice);
alpha = 1.0;
alpham1 = -1.0;
beta = 0.0;
r0 = 0.;
printf("\n Data transferred\n");
cudaEvent_t start,stop;
cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord(start, 0);
cusparseDcsrmv(cusparseHandle,CUSPARSE_OPERATION_NON_TRANSPOSE, (m+n), (m+n), G.nonzero, &alpha, descr, d_val, d_row, d_col, d_x, &beta, d_Ax);
cublasDaxpy(cublasHandle, (m+n), &alpham1, d_Ax, 1, d_r, 1);
cublasStatus = cublasDdot(cublasHandle, (m+n), d_r, 1, d_r, 1, &r1);
k = 1;
while (r1 > tol && k <= max_iter)
{
if (k > 1)
{
b = r1 / r0;
cublasStatus = cublasDscal(cublasHandle, (m+n), &b, d_p, 1);
cublasStatus = cublasDaxpy(cublasHandle, (m+n), &alpha, d_r, 1, d_p, 1);
}
else
{
cublasStatus = cublasDcopy(cublasHandle, (m+n), d_r, 1, d_p, 1);
}
cusparseDcsrmv(cusparseHandle, CUSPARSE_OPERATION_NON_TRANSPOSE, (m+n), (m+n), G.nonzero, &alpha, descr, d_val, d_row, d_col, d_p, &beta, d_Ax);
cublasStatus = cublasDdot(cublasHandle, (m+n), d_p, 1, d_Ax, 1, &dot);
a = r1 / dot;
cublasStatus = cublasDaxpy(cublasHandle, (m+n), &a, d_p, 1, d_x, 1);
na = -a;
cublasStatus = cublasDaxpy(cublasHandle, (m+n), &na, d_Ax, 1, d_r, 1);
r0 = r1;
cublasStatus = cublasDdot(cublasHandle, (m+n), d_r, 1, d_r, 1, &r1);
// cudaThreadSynchronize();
// printf("iteration = %3d, residual = %e\n", k, sqrt(r1));
k++;
}
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
float elapsedTime;
cudaEventElapsedTime(&elapsedTime, start, stop);
printf("Iterations = %3d\tTime : %.6f milli-seconds : \n", k, elapsedTime);
cudaMemcpy(G.x, d_x, (m+n)*sizeof(double), cudaMemcpyDeviceToHost);
/* printf("\n x = \n");
for(i=0;i<(m+n);i++)
{
printf("\n x[%d] = %.8f", i, G.x[i]);
}
*/ float rsum, diff, err = 0.0;
for (int i = 0; i < (m+n); i++)
{
rsum = 0.0;
for (int j = G.rowIndex[i]; j < G.rowIndex[i+1]; j++)
{
rsum += G.Mat_val[j]*G.x[G.columns[j]];
}
diff = fabs(rsum - G.b[i]);
if (diff > err)
{
err = diff;
}
}
cusparseDestroy(cusparseHandle);
cublasDestroy(cublasHandle);
/* free(I);
free(J);
free(val);
free(x);
free(rhs);
*/ cudaFree(d_col);
cudaFree(d_row);
cudaFree(d_val);
cudaFree(d_x);
cudaFree(d_r);
cudaFree(d_p);
cudaFree(d_Ax);
// cudaDeviceReset causes the driver to clean up all state. While
// not mandatory in normal operation, it is good practice. It is also
// needed to ensure correct operation when the application is being
// profiled. Calling cudaDeviceReset causes all profile data to be
// flushed before the application exits
cudaDeviceReset();
/* printf("\n X is:\n");
for(i=0;i<(m+n);i++)
{
printf("\n X[%d] = %.4f", i, G.x[i]);
}
*/
printf("Test Summary: Error amount = %f\n", err);
exit((k <= max_iter) ? 0 : 1);
/*
culaSparseHandle handle;
if (culaSparseCreate(&handle) != culaSparseNoError)
{
// this should only fail under extreme conditions
std::cout << "fatal error: failed to create library handle!" << std::endl;
exit(EXIT_FAILURE);
}
StatusChecker sc(handle);
culaSparsePlan plan;
sc = culaSparseCreatePlan(handle, &plan);
culaSparseCsrOptions formatOpts;
culaSparseCsrOptionsInit(handle, &formatOpts);
//formatOpts.indexing=1;
sc = culaSparseSetDcsrData(handle, plan, &formatOpts, m+n, G.nonzero, &G.Mat_val[0], &G.rowIndex[0], &G.columns[0], &G.x[0], &G.b[0]);
printf("\n Fine till here");
printf("\n");
culaSparseConfig config;
sc = culaSparseConfigInit(handle, &config);
config.relativeTolerance = 1e-2;
config.maxIterations = 100000;
config.divergenceTolerance = 20;
culaSparseResult result;
/* sc = culaSparseSetHostPlatform(handle, plan, 0);
// set cg solver
sc = culaSparseSetCgSolver(handle, plan, 0);
printf("\n Fine till here");
printf("\n");
// perform solve (cg + no preconditioner on host)
culaSparseResult result;
sc = culaSparseExecutePlan(handle, plan, &config, &result);
sc = culaSparseGetResultString(handle, &result, buffer, bufsize);
std::cout << buffer << std::endl;
if (culaSparsePreinitializeCuda(handle) == culaSparseNoError)
{
// change to cuda accelerated platform
sc = culaSparseSetCudaPlatform(handle, plan, 0);
// perform solve (cg + ilu0 on cuda)
culaSparseGmresOptions solverOpts;
culaSparseStatus status = culaSparseGmresOptionsInit(handle, &solverOpts);
solverOpts.restart = 20;
sc = culaSparseSetCgSolver(handle, plan, 0);
// sc = culaSparseSetJacobiPreconditioner(handle, plan, 0);
// sc = culaSparseSetIlu0Preconditioner(handle, plan, 0);
sc = culaSparseExecutePlan(handle, plan, &config, &result);
sc = culaSparseGetResultString(handle, &result, buffer, bufsize);
std::cout << buffer << std::endl;
// change preconditioner to fainv
// this avoids data transfer by using data cached by the plan
// sc = culaSparseSetIlu0Preconditioner(handle, plan, 0);
// perform solve (cg + fainv on cuda)
// these timing results should indicate minimal overhead
/* sc = culaSparseExecutePlan(handle, plan, &config, &result);
sc = culaSparseGetResultString(handle, &result, buffer, bufsize);
std::cout << buffer << std::endl;
// change solver
// this avoids preconditioner regeneration by using data cached by the plan
sc = culaSparseSetBicgstabSolver(handle, plan, 0);
// perform solve (bicgstab + fainv on cuda)
// the timing results should indicate minimal overhead and preconditioner generation time
sc = culaSparseExecutePlan(handle, plan, &config, &result);
sc = culaSparseGetResultString(handle, &result, buffer, bufsize);
std::cout << buffer << std::endl;
sc = culaSparseSetGmresSolver(handle, plan, 0);
// perform solve (bicgstab + fainv on cuda)
// the timing results should indicate minimal overhead and preconditioner generation time
sc = culaSparseExecutePlan(handle, plan, &config, &result);
sc = culaSparseGetResultString(handle, &result, buffer, bufsize);
std::cout << buffer << std::endl;
}
else
{
std::cout << "alert: no cuda capable gpu found" << std::endl;
}
// cleanup plan
culaSparseDestroyPlan(plan);
// cleanup handle
culaSparseDestroy(handle);
FILE* myWriteFile;
myWriteFile=fopen("result.txt","w");
for (i = 0; i < n; i++)
{
fprintf(myWriteFile,"%1f\n",G.x[i]);
// printf ("\n x [%d] = % f", i, x[i]);
}
fprintf(myWriteFile,".end\n");
fclose(myWriteFile);
printf ("\n");
// time_st=dsecnd();
// solver(G.rowIndex,G.columns,G.Mat_val,G.b,G.x,m+n,G.nonzero);
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// printf("Successfully Solved in : %.6f secs\n",time_avg);
cerr<<endl;
cerr<<"Fillup Graph ";
// time_st=dsecnd();
G.fillup_graph();
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cerr<<"Done "<<time_avg<<endl;
//G.output_graph_stdout();
cerr<<"Matching KCL ";
// time_st=dsecnd();
G.check_kcl();
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cerr<<"Done "<<time_avg<<endl;
/*for (int i=0;i<m+n;i++)
{
cout<<"M"<<i<<endl;
for (int j=0;j<m+n;j++)
cout<<" "<<j<<"#"<<M[i][j]<<endl;
}*/
}
magma_int_t
magma_d_spmv( double alpha, magma_d_sparse_matrix A,
magma_d_vector x, double beta, magma_d_vector y )
{
if( A.memory_location != x.memory_location ||
x.memory_location != y.memory_location ){
printf("error: linear algebra objects are not located in same memory!\n");
printf("memory locations are: %d %d %d\n",
A.memory_location, x.memory_location, y.memory_location );
return MAGMA_ERR_INVALID_PTR;
}
// DEV case
if( A.memory_location == Magma_DEV ){
if( A.num_cols == x.num_rows ){
if( A.storage_type == Magma_CSR
|| A.storage_type == Magma_CSRL
|| A.storage_type == Magma_CSRU ){
//printf("using CSR kernel for SpMV: ");
//magma_dgecsrmv( MagmaNoTrans, A.num_rows, A.num_cols, alpha,
// A.val, A.row, A.col, x.val, beta, y.val );
//printf("done.\n");
cusparseHandle_t cusparseHandle = 0;
cusparseStatus_t cusparseStatus;
cusparseStatus = cusparseCreate(&cusparseHandle);
cusparseMatDescr_t descr = 0;
cusparseStatus = cusparseCreateMatDescr(&descr);
cusparseSetMatType(descr,CUSPARSE_MATRIX_TYPE_GENERAL);
cusparseSetMatIndexBase(descr,CUSPARSE_INDEX_BASE_ZERO);
cusparseDcsrmv( cusparseHandle,CUSPARSE_OPERATION_NON_TRANSPOSE,
A.num_rows, A.num_cols, A.nnz, &alpha, descr,
A.val, A.row, A.col, x.val, &beta, y.val );
cusparseDestroyMatDescr( descr );
cusparseDestroy( cusparseHandle );
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_ELLPACK ){
//printf("using ELLPACK kernel for SpMV: ");
magma_dgeellmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.val, A.col, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_ELL ){
//printf("using ELL kernel for SpMV: ");
magma_dgeelltmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.val, A.col, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_ELLRT ){
//printf("using ELLRT kernel for SpMV: ");
magma_dgeellrtmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.max_nnz_row, alpha, A.val, A.col, A.row, x.val,
beta, y.val, A.alignment, A.blocksize );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_SELLC ){
//printf("using SELLC kernel for SpMV: ");
magma_dgesellcmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.blocksize, A.numblocks, A.alignment,
alpha, A.val, A.col, A.row, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_SELLP ){
//printf("using SELLP kernel for SpMV: ");
magma_dgesellpmv( MagmaNoTrans, A.num_rows, A.num_cols,
A.blocksize, A.numblocks, A.alignment,
alpha, A.val, A.col, A.row, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_DENSE ){
//printf("using DENSE kernel for SpMV: ");
magmablas_dgemv( MagmaNoTrans, A.num_rows, A.num_cols, alpha,
A.val, A.num_rows, x.val, 1, beta, y.val, 1 );
//printf("done.\n");
return MAGMA_SUCCESS;
}
/* else if( A.storage_type == Magma_BCSR ){
//printf("using CUSPARSE BCSR kernel for SpMV: ");
// CUSPARSE context //
cusparseHandle_t cusparseHandle = 0;
cusparseStatus_t cusparseStatus;
cusparseStatus = cusparseCreate(&cusparseHandle);
cusparseMatDescr_t descr = 0;
cusparseStatus = cusparseCreateMatDescr(&descr);
// end CUSPARSE context //
cusparseDirection_t dirA = CUSPARSE_DIRECTION_ROW;
int mb = (A.num_rows + A.blocksize-1)/A.blocksize;
int nb = (A.num_cols + A.blocksize-1)/A.blocksize;
cusparseDbsrmv( cusparseHandle, dirA,
CUSPARSE_OPERATION_NON_TRANSPOSE, mb, nb, A.numblocks,
&alpha, descr, A.val, A.row, A.col, A.blocksize, x.val,
&beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}*/
else {
printf("error: format not supported.\n");
return MAGMA_ERR_NOT_SUPPORTED;
}
}
else if( A.num_cols < x.num_rows ){
magma_int_t num_vecs = x.num_rows / A.num_cols;
if( A.storage_type == Magma_CSR ){
//printf("using CSR kernel for SpMV: ");
magma_dmgecsrmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, alpha, A.val, A.row, A.col, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_ELLPACK ){
//printf("using ELLPACK kernel for SpMV: ");
magma_dmgeellmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, A.max_nnz_row, alpha, A.val, A.col, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}
else if( A.storage_type == Magma_ELL ){
//printf("using ELL kernel for SpMV: ");
magma_dmgeelltmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, A.max_nnz_row, alpha, A.val,
A.col, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}else if( A.storage_type == Magma_SELLP ){
//printf("using SELLP kernel for SpMV: ");
magma_dmgesellpmv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, A.blocksize, A.numblocks, A.alignment,
alpha, A.val, A.col, A.row, x.val, beta, y.val );
//printf("done.\n");
return MAGMA_SUCCESS;
}/*
if( A.storage_type == Magma_DENSE ){
//printf("using DENSE kernel for SpMV: ");
magmablas_dmgemv( MagmaNoTrans, A.num_rows, A.num_cols,
num_vecs, alpha, A.val, A.num_rows, x.val, 1,
beta, y.val, 1 );
//printf("done.\n");
return MAGMA_SUCCESS;
}*/
else {
printf("error: format not supported.\n");
return MAGMA_ERR_NOT_SUPPORTED;
}
}
}
// CPU case missing!
else{
printf("error: CPU not yet supported.\n");
return MAGMA_ERR_NOT_SUPPORTED;
}
return MAGMA_SUCCESS;
}
void testCusolver(int rows, int cols, int nnz, int *row_ptr, int *col_index, double *values,
double *valuesB){
// --- Initialize cuSPARSE
cusolverSpHandle_t cusolver_handle = NULL; checkCudaErrors(cusolverSpCreate(&cusolver_handle));
cusparseHandle_t cusparse_handle = NULL; checkCudaErrors(cusparseCreate(&cusparse_handle));
cudaStream_t cudaStream = NULL; checkCudaErrors(cudaStreamCreate(&cudaStream));
checkCudaErrors(cusolverSpSetStream(cusolver_handle, cudaStream));
checkCudaErrors(cusparseSetStream(cusparse_handle, cudaStream));
cusparseMatDescr_t descrA; checkCudaErrors(cusparseCreateMatDescr(&descrA));
checkCudaErrors(cusparseSetMatType (descrA, CUSPARSE_MATRIX_TYPE_GENERAL));
checkCudaErrors(cusparseSetMatIndexBase(descrA, CUSPARSE_INDEX_BASE_ONE));
double start, stop, time_to_solve;
start = second();
// --- Device side dense matrix
printf("\nAlloc GPU memory...\n");
double *d_A; checkCudaErrors(cudaMalloc(&d_A, nnz * sizeof(double)));
int *d_A_RowIndices; checkCudaErrors(cudaMalloc(&d_A_RowIndices, (rows + 1) * sizeof(int)));
int *d_A_ColIndices; checkCudaErrors(cudaMalloc(&d_A_ColIndices, nnz * sizeof(int)));
double *d_x; checkCudaErrors(cudaMalloc(&d_x, rows * sizeof(double)));
checkCudaErrors(cudaMemcpy(d_A, values, nnz * sizeof(double), cudaMemcpyHostToDevice));
checkCudaErrors(cudaMemcpy(d_A_RowIndices, row_ptr, (rows + 1) * sizeof(int), cudaMemcpyHostToDevice));
checkCudaErrors(cudaMemcpy(d_A_ColIndices, col_index, nnz * sizeof(int), cudaMemcpyHostToDevice));
double *d_b; checkCudaErrors(cudaMalloc(&d_b, rows * sizeof(double)));
checkCudaErrors(cudaMemcpy(d_b, valuesB, rows * sizeof(double), cudaMemcpyHostToDevice));
double *h_x = (double *)malloc(rows * sizeof(double));
double tol = 1.e-12;
int reorder = 0;
int singularity = 0;
printf("\nProcessing in GPU using cusolver QR...\n");
//checkCudaErrors(cusolverSpDcsrlsvluHost(cusolver_handle, Nrows, nnz, descrA, sparse.Values(),
// sparse.RowPtr(), sparse.ColIdx(), mB.values, tol, reorder, h_x, &singularity));
checkCudaErrors(cusolverSpDcsrlsvqr(cusolver_handle, rows, nnz, descrA, d_A, d_A_RowIndices,
d_A_ColIndices, d_b, tol, reorder, d_x, &singularity));
checkCudaErrors(cudaDeviceSynchronize());
stop = second();
time_to_solve = stop - start;
checkCudaErrors(cudaMemcpy(h_x, d_x, rows * sizeof(double), cudaMemcpyDeviceToHost));
double minusOne = -1.0;
double one = 1.0;
double *d_r; checkCudaErrors(cudaMalloc((void **)&d_r, sizeof(double)*rows));
checkCudaErrors(cudaMemcpy(d_r, d_b, sizeof(double)*rows, cudaMemcpyDeviceToDevice));
checkCudaErrors(cusparseDcsrmv(cusparse_handle,
CUSPARSE_OPERATION_NON_TRANSPOSE,
rows,
cols,
nnz,
&minusOne,
descrA,
d_A,
d_A_RowIndices,
d_A_ColIndices,
d_x,
&one,
d_r));
double *h_r; h_r = (double*) malloc(rows * sizeof(double));
checkCudaErrors(cudaMemcpy(h_r, d_r, sizeof(double)*rows, cudaMemcpyDeviceToHost));
checkCudaErrors(cudaMemcpy(h_r, d_r, rows * sizeof(double), cudaMemcpyDeviceToHost));
double r_inf = vec_norminf(rows, h_r);
printf("(GPU - cuSolver) Time (sec): %f\n", time_to_solve);
printf("(Eigen) |b - A*x| = %E \n", r_inf);
checkCudaErrors(cusparseDestroy(cusparse_handle));
checkCudaErrors(cusolverSpDestroy(cusolver_handle));
checkCudaErrors(cudaStreamDestroy(cudaStream));
checkCudaErrors(cudaFree(d_b));
checkCudaErrors(cudaFree(d_x));
checkCudaErrors(cudaFree(d_r));
checkCudaErrors(cudaFree(d_A));
checkCudaErrors(cudaFree(d_A_RowIndices));
checkCudaErrors(cudaFree(d_A_ColIndices));
free(h_x);
free(h_r);
}
int main(int argc, char* argv[])
{
const int bufsize = 512;
char buffer[bufsize];
int m,n,S;
double time_st,time_end,time_avg;
//omp_set_num_threads(2);
// printf("\n-----------------\nnumber of threads fired = %d\n-----------------\n",(int)omp_get_num_threads());
if(argc!=2)
{
cout<<"Insufficient arguments"<<endl;
return 1;
}
graph G;
cerr<<"Start reading ";
// time_st=dsecnd();
G.create_graph(argv[1]);
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cout<<"Success "<<endl;
// cerr<<"Reading time "<<time_avg<<endl;
cerr<<"Constructing Matrices ";
// time_st=dsecnd();
G.construct_MNA();
G.construct_NA();
// time_end=dsecnd();
// time_avg = (time_end-time_st);
// cerr<<"Done "<<time_avg<<endl;
// G.construct_sparse_MNA();
m=G.node_array.size()-1;
n=G.voltage_edge_id.size();
cout<<endl;
cout<<"MATRIX STAT:"<<endl;
cout<<"Nonzero elements: "<<G.nonzero<<endl;
cout<<"Number of Rows: "<<m+n<<endl;
cout<<"Nonzero in G: "<<G.Gnonzero<<endl;
cout<<"Number of rows in G: "<<m<<endl;
cout<<"Nonzero in P: "<<G.Pnonzero<<endl;
cout<<"Number of rows in P: "<<m<<endl;
// printf("\n Nonzero = %d", G.nonzero);
// printf("\n Rows = %d", m+n);
cout<<"MAT val: "<<endl;
int i,j;
G.Mat_val[0] += 100;
G.Gmat[0] +=100;
/*
for(i=0;i<G.Gnonzero;i++)
cout<<" "<<G.Gmat[i];
cout<<endl;
for(i=0;i<G.Gnonzero;i++)
cout<<" "<<G.Gcolumns[i];
cout<<endl;
for(i=0;i<m+1;i++)
cout<<" "<<G.GrowIndex[i];
cout<<endl;
for(i=0;i<m;i++)
printf(" %.8f", G.b1[i]);
cout<<endl;
for(i=0;i<m;i++)
printf(" %.8f", G.x1[i]);
cout<<endl;
*/ SuiteSparse_long *Gnz = (SuiteSparse_long*)calloc(m,sizeof(SuiteSparse_long));
for(i=0;i<m;i++)
{
// cout<<endl;
SuiteSparse_long startindex=G.GrowIndex[i];
SuiteSparse_long endindex=G.GrowIndex[i+1];
Gnz[i] = endindex - startindex;
// for(j=startindex;j<endindex;j++)
// cout<<" "<<G.Gmat[j];
// cout<<endl;
}
/* for(i=0;i<G.Pnonzero;i++)
cout<<" "<<G.Pmat[i];
cout<<endl;
for(i=0;i<G.Pnonzero;i++)
cout<<" "<<G.Pcolumns[i];
cout<<endl;
for(i=0;i<m+1;i++)
cout<<" "<<G.ProwIndex[i];
cout<<endl;
/* for(i=0;i<m;i++)
printf(" %.8f", G.b1[i]);
cout<<endl;
for(i=0;i<m;i++)
printf(" %.8f", G.x1[i]);
cout<<endl;
for(i=0;i<m;i++)
{
cout<<endl;
int startindex=G.ProwIndex[i];
int endindex=G.ProwIndex[i+1];
for(j=startindex;j<endindex;j++)
cout<<" "<<G.Pmat[j];
cout<<endl;
}
/* for(i=0;i<G.nonzero;i++)
cout<<" "<<G.Mat_val[i];
cout<<endl;
for(i=0;i<G.nonzero;i++)
cout<<" "<<G.columns[i];
cout<<endl;
for(i=0;i<m+n+1;i++)
cout<<" "<<G.rowIndex[i];
cout<<endl;
for(i=0;i<m+n;i++)
printf(" %.8f", G.b[i]);
cout<<endl;
for(i=0;i<m+n;i++)
printf(" %.8f", G.x[i]);
cout<<endl;
for(i=0;i<m+n;i++)
{
cout<<endl;
int startindex=G.rowIndex[i];
int endindex=G.rowIndex[i+1];
for(j=startindex;j<endindex;j++)
cout<<" "<<G.Mat_val[j];
cout<<endl;
}
*/
/* for (i=0;i<m+n+1;i++)
{
//cout<<endl;
if(G.rowIndex[i]==G.rowIndex[i+1])
break;
for(j=G.rowIndex[i];j<G.rowIndex[i+1];j++)
{
if(G.Mat_val[j]>10)
cout<<G.Mat_val[j]<<"\t";
}
//cout<<endl;
/*for(j=G.rowIndex[i];j<G.rowIndex[i+1];j++)
{
cout<<G.columns[j]<<"\t";
}
//cout<<endl;
}
cout<<endl;
*/
//printing the matrix
printf("\n Fine till here");
printf("\n");
// int* rowmIndex=(int*)calloc(m+1,sizeof(int));
printf("\n Fine till here");
printf("\n");
//int rowmIndex[5]={1,2,3,4,5};
/* for(i=0;i<m+1;i++)
{
rowmIndex[i]=G.rowIndex[i];
printf(" %d", rowmIndex[i]);
}
*/
printf("\n Allocating GPU memory\n");
cudaDeviceReset();
size_t free, total;
cudaMemGetInfo(&free, &total);
printf("\n Free Mem = %lf MB, Total mem = %lf MB\n", (double)(free)/(1024*1024), (double)(total)/(1024*1024));
double *dev_csrValA, *dev_b, *dev_x;
int *dev_csrRowIdxA, *dev_csrColA;
double *dev_GcsrVal, *dev_b1, *dev_x1;
double *dev_PcsrVal, *dev_b2, *dev_x2;
int *dev_GcsrRowIdx, *dev_PcsrRowIdx, *dev_GcsrCol, *dev_PcsrCol;
cudaMalloc((void**)&dev_PcsrVal, G.Pnonzero*sizeof(double));
cudaMalloc((void**)&dev_PcsrRowIdx, (m+1)*sizeof(int));
cudaMalloc((void**)&dev_PcsrCol, G.Pnonzero*sizeof(int));
cudaMalloc((void**)&dev_b1, (m)*sizeof(double));
cudaMalloc((void**)&dev_b2, n*sizeof(double));
cudaMalloc((void**)&dev_x1, m*sizeof(double));
cudaMalloc((void**)&dev_x2, n*sizeof(double));
cudaMemcpy(dev_b1, G.b1, (m)*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(dev_x1, G.x1, (m)*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(dev_PcsrVal, G.Pmat, G.Pnonzero*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(dev_b2, G.b2, (n)*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(dev_x2, G.x2, (n)*sizeof(double), cudaMemcpyHostToDevice);
cudaMemcpy(dev_PcsrRowIdx, G.ProwIndex, (m+1)*sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_PcsrCol, G.Pcolumns, (G.Pnonzero)*sizeof(int), cudaMemcpyHostToDevice);
/* Matrix has been created and stored in CSR format.
However, CHOLMOD requires CSC format. Since our matrix is symmetric positive definite, we can simply swap
csrColA with csrRowIdx and vice versa
*/
/* Starting the CHOLMOD routine now*/
printf("\n Initiating CHOLMOD\n");
cholmod_sparse *A, *P;
cholmod_dense *x, *b, *r, *midvec;
cholmod_factor *L;
cholmod_common *Common, cm;
Common = &cm;
cholmod_l_start(Common);
// &Common->useGPU=1;
printf("\n m = %d, G.Gnonzero = %d\n", m, G.Gnonzero);
cholmod_sparse *C = cholmod_l_allocate_sparse((size_t)(m), (size_t)(m), (size_t)(G.Gnonzero), 1, 0, 1, 1, Common);
// P = cholmod_l_allocate_sparse((size_t)(m), (size_t)(n), (size_t)(G.Pnonzero), 1, 0, 0, 1, Common);
// printf("\n Allocated \n");
C->itype = CHOLMOD_LONG;
// printf("\n Itype \n");
C->p = &G.GrowIndex[0];
// printf("\n Columns \n");
C->nz = &Gnz[0];
// printf("\n Rows \n");
C->i = &G.Gcolumns[0];
C->dtype = 0;
C->x = &G.Gmat[0];
/* P->itype = CHOLMOD_LONG;
P->p = &G.ProwIndex[0];
P->nz = &Pnz[0];
P->i = &G.Pcolumns[0];
P->dtype = 0;
P->x = &G.Pmat[0];
*/
b = cholmod_l_allocate_dense((size_t)(m), 1, (size_t)(m), 1, Common);
b->dtype=0;
b->x = &G.b1[0];
b->xtype = 1;
printf("\n CHOLMOD manually set\n");
cholmod_l_print_sparse(C, "A", Common);
cholmod_l_print_dense(b, "b", Common);
cudaEvent_t start, stop;
cudaEventCreate(&start);
cudaEventCreate(&stop);
cudaEventRecord(start, 0);
L = cholmod_l_analyze(C, Common);
printf("\n Analysis: Flops: %g \t lnz: %g\n", Common->fl, Common->lnz);
cholmod_l_factorize(C, L, Common);
x = cholmod_l_solve(CHOLMOD_A, L, b, Common);
cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
float elapsedTime;
cudaEventElapsedTime(&elapsedTime, start, stop);
printf("\n Time : %.6f secs :\n", elapsedTime);
cholmod_l_print_dense(x, "X", Common);
double *x1_mod = (double*)x->x;
cudaMemcpy(dev_x1, x1_mod, m*sizeof(double), cudaMemcpyHostToDevice);
cusparseStatus_t cuSparseStatus;
cusparseHandle_t cuspHandle;
cuSparseStatus = cusparseCreate(&cuspHandle);
cusparseMatDescr_t descrP;
cusparseCreateMatDescr(&descrP);
cusparseSetMatType(descrP, CUSPARSE_MATRIX_TYPE_GENERAL);
cusparseSetMatIndexBase(descrP, CUSPARSE_INDEX_BASE_ZERO);
double *dev_res1, *dev_simple;
double *res1 = (double*)calloc(n,sizeof(double));
cudaMalloc((void**)&dev_res1, n*sizeof(double));
cudaMalloc((void**)&dev_simple, n*sizeof(double));
const double alpha = 1.0, beta=0.0;
//alpha = 1.0;
//beta = 0.0;
//solving P^T * G^-1 * b1 Result stored in dev_res1
cuSparseStatus = cusparseDcsrmv(cuspHandle, CUSPARSE_OPERATION_TRANSPOSE, m, n, G.Pnonzero, &alpha, descrP, dev_PcsrVal, dev_PcsrRowIdx, dev_PcsrCol, dev_x1, &beta, dev_res1);
if(cuSparseStatus == CUSPARSE_STATUS_SUCCESS)
{
/* cudaMemcpy(res1, dev_res1, n*sizeof(double), cudaMemcpyDeviceToHost);
for(i=0;i<n;i++)
{
printf("\nres1[%d] = %.8f", i, res1[i]);
}
printf("\n P^T * G^-1 * b1 done! Vector stored in res1");
*/ }
else
{
printf("\n P^T * G^-1 * b1 failed\n");
exit(1);
}
const double alphaneg = -1.0;
//Solving P^T * G^-1 * b1 - b2 ; Result stored in dev_res1
cublasStatus_t cuBlasStatus;
cublasHandle_t cubHandle;
cuBlasStatus = cublasCreate(&cubHandle);
cuBlasStatus = cublasDaxpy(cubHandle, n, &alphaneg, dev_b2, 1, dev_res1, 1);
if(cuBlasStatus == CUBLAS_STATUS_SUCCESS)
{
// cudaMemcpy(res1, dev_res1, n*sizeof(double), cudaMemcpyDeviceToHost);
// for(i=0;i<n;i++)
// {
// printf("\nres1[%d] = %.8f", i, res1[i]);
// }
printf("\n res1 = res1 - b2 done\n");
}
else
{
printf("\n res1 = res1 - b2 failed\n");
}
///NOW COMPUTING G^-1 * P
int k = 0;
int breakloop=0;
double **midMat = (double**)malloc(m*sizeof(double*));
for(i=0;i<m;i++)
{
midMat[i] = (double*)calloc(n,sizeof(double));
}
cudaEventRecord(start, 0);
for(i=0;i<n;i++)
{
breakloop = 0;
double *vect = (double*)calloc(m,sizeof(double*));
for(j=0;j<m;j++)
{
int startin = G.ProwIndex[j];
int endin = G.ProwIndex[j+1];
if(startin == endin)
continue;
k = startin;
while(k<endin)
{
if(G.Pcolumns[k] == i)
{
vect[j] = G.Pmat[k];
breakloop=1;
break;
}
k++;
}
if(breakloop == 1)
{
break;
}
}
midvec = cholmod_l_allocate_dense((size_t)(m), 1, (size_t)(m), 1, Common);
midvec->dtype=0;
midvec->x=&vect[0];
midvec->xtype = 1;
cholmod_dense *res2;
res2 = cholmod_l_solve(CHOLMOD_A, L, midvec, Common);
double *re = (double*)res2->x;
// printf("\n vector %d is:\n", i);
int i1, j1, k1;
// for(j1=0;j1<m;j1++)
// {
// midmat2flat[i+j1*n] = re[j1];
// printf(" %lf", re[j1]);
// }
// printf("\n");
for(i1=0;i1<m;i1++)
{
midMat[i1][i] = re[i1];
}
cholmod_l_free_dense(&midvec, Common);
}
/* printf("\n Midmat = \n");
for(i=0;i<m;i++)
{
for(j=0;j<n;j++)
{
printf(" %lf", midMat[i][j]);
}
printf("\n");
}
*/
double *midMatflat = (double*)calloc((m*n),sizeof(double));
double *dev_midMat;
double *dev_solut;
int counter = 0;
for(i=0;i<n;i++)
{
for(j=0;j<m;j++)
{
midMatflat[counter] = midMat[j][i];
counter++;
}
}
cudaMalloc((void**)&dev_midMat, m*n*sizeof(double));
cudaMalloc((void**)&dev_solut, n*n*sizeof(double));
cudaMemcpy(dev_midMat, midMatflat, m*n*sizeof(double), cudaMemcpyHostToDevice);
//Solving P^T * midMat; Result stored in dev_solut
cuSparseStatus = cusparseDcsrmm(cuspHandle, CUSPARSE_OPERATION_TRANSPOSE, m, n, n, G.Pnonzero, &alpha, descrP, dev_PcsrVal, dev_PcsrRowIdx, dev_PcsrCol, dev_midMat, m, &beta, dev_solut, n);
if(cuSparseStatus == CUSPARSE_STATUS_SUCCESS)
{
printf("\n Solved P^T * G^-1 * P. Result stored in solut\n");
}
else
{
printf("\n Failed to Solve P^T * G^-1 * P \n");
exit(1);
}
/* double *matGflat = (double*)calloc(n*n,sizeof(double));
cudaMemcpy(matGflat, dev_solut, n*n*sizeof(double), cudaMemcpyDeviceToHost);
counter = 0;
printf("\nBefore LU starts\n");
for(i=0;i<n;i++)
{
for(j=0;j<n;j++)
{
printf(" %lf ", matGflat[counter]);
counter++;
}
printf("\n");
}
printf("\n");
*/
cusolverStatus_t cuSolverStatus;
cusolverDnHandle_t cudenHandle;
cuSolverStatus = cusolverDnCreate(&cudenHandle);
int Lwork = 0;
cuSolverStatus = cusolverDnDgetrf_bufferSize(cudenHandle, n, n, dev_solut, n, &Lwork);
if(cuSolverStatus == CUSOLVER_STATUS_SUCCESS)
{
printf("\n Buffer works\n Lwork = %d\n", Lwork);
}
else
{
exit(1);
}
double *dev_Workspace;
int *dev_Ipiv, *dev_Info;
cudaMalloc((void**)&dev_Workspace, Lwork*sizeof(double));
cudaMalloc((void**)&dev_Ipiv, n*sizeof(int));
cudaMalloc((void**)&dev_Info, sizeof(int));
//Calculating LU for dev_solut
// double *nnmat = (double*)calloc(n*n,sizeof(double));
// cudaMemcpy(nnmat, dev_solut, n*n*sizeof(double), cudaMemcpyDeviceToHost);
// cuSolverStatus = cusolverDnDgetrfHost(cudenHandle, n, n,
cuSolverStatus = cusolverDnDgetrf(cudenHandle, n, n, dev_solut, n, dev_Workspace, dev_Ipiv, dev_Info);
if(cuSolverStatus == CUSOLVER_STATUS_SUCCESS)
{
printf("\n solut has be defactorized into L and U. dev_Ipiv * solut = L * U\n");
}
else
{
printf("\n Unable to defactorize solut into LU\n");
exit(1);
}
//solving dev_solut * x = dev_res1. Result stored in dev_res1
cuSolverStatus = cusolverDnDgetrs(cudenHandle, CUBLAS_OP_N, n, 1, dev_solut, n, dev_Ipiv, dev_res1, n, dev_Info);
if(cuSolverStatus == CUSOLVER_STATUS_SUCCESS)
{
printf("\n Solution obtained for x2 \n");
}
else
{
printf("\n LU decomposition obtained by LU solver failed\n");
}
/* cudaMemcpy(G.x2, dev_res1, n*sizeof(double), cudaMemcpyDeviceToHost);
printf("\n x2 = \n");
for(i=0;i<n;i++)
{
printf("\n x2[%d] = %lf", i, G.x2[i]);
}
*/
double *dev_dummy;
cudaMalloc((void**)&dev_dummy, m*sizeof(double));
cudaMemset(dev_dummy, 0.0, m*sizeof(double));
printf("\n Starting solving for x1 \n");
//Solving for x1
//Solving G^-1 * P * x2; G^-1 * P is stored in midMat
cuBlasStatus = cublasDgemv(cubHandle, CUBLAS_OP_N, m, n, &alpha, dev_midMat, m, dev_res1, 1, &beta, dev_dummy, 1);
if(cuBlasStatus == CUBLAS_STATUS_SUCCESS)
{
/* double *toprint = (double*)calloc(m,sizeof(double));
cudaMemcpy(toprint, dev_dummy, m*sizeof(double), cudaMemcpyDeviceToHost);
printf("\n Intermediate vector :\n");
for(i=0;i<m;i++)
{
printf("\ndummy[%d] = %lf", i, toprint[i]);
}
*/ printf("\n midmat * x2 obtained. Stored in dummy\n");
}
else
{
printf("\n Failed to obtain midmat * x2\n");
}
cuBlasStatus = cublasDaxpy(cubHandle, m, &alphaneg, dev_dummy, 1, dev_x1, 1);
if(cuBlasStatus == CUBLAS_STATUS_SUCCESS)
{
/* cudaMemcpy(G.x1, dev_x1, m*sizeof(double), cudaMemcpyDeviceToHost);
printf("\n x1 = \n");
for(i=0;i<m;i++)
{
printf("\n x1[%d] = %.15f", i, G.x1[i]);
}
*/ printf("\n x1 obtained");
}
else
{
printf("\n Failed to obtain x1");
}
printf("\n Solver finished its work\n");
/* cudaEventRecord(stop, 0);
cudaEventSynchronize(stop);
cudaEventElapsedTime(&elapsedTime, start, stop);
printf("\n Time: %.6f msecs :\n", elapsedTime);
*/
cholmod_l_finish(Common);
return 0;
}
