how to avoid copying to/from the GPU

[vish108]

Hi All,

I am having to call magma functions within a loop. There are dependencies between the function calls, so I have to copy the result from the previous call to the host, then call the next function, and so on. Consequently, the code spends most of its time in transferring data to/from the GPU.

Is there a way to minimize data transfer in such a scenario?

Sincerely,

Vishal

[hsahasra]

Yes, transfer the starting data once then successively use magma_*_gpu calls. _gpu calls take GPU data pointers. You can also have a look at CUBLAS and CUSPARSE libraries.

http://docs.nvidia.com/cuda/cublas/index.html
http://docs.nvidia.com/cuda/cusparse/

[mgates3]

Can you give a bit more detail about what you are doing? Some pseudocode, giving what routines you are calling, would be helpful. It sounds like you are doing something like this:

Code: Select all

while( condition ) {
    magma_dgesv_gpu( ..., dA, ..., dB, ... );
    magma_dgetmatrix( ..., dA, ..., A, ... );
}

You might consider using the CPU interfaces, for example, magma_dgesv, instead of GPU interfaces, magma_dgesv_gpu. The CPU interface copies the matrix to the GPU, overlapping the copy with some computation to reduce its cost.

-mark

[vish108]

Hi Mark,

Sure, here is a part of the pseudocode. My program uses multiple threads, and based on my understanding MAGMA uses cuda streams which allow it to launch multiple kernels in parallel. In f1(), I copy the matrix to the GPU calculate the inverse, copy the inverse back to the CPU. Then, I call another function to perform scalar multiplication on the inverted matrix. As you will see, I am doing this naively (am very new to GPU programming). Am sure that there must be a way to avoid the multiple data transfers. Please advise:

Code: Select all

#pragma omp parallel for schedule(static,1) num_threads(8)
for(int i = 0; i < 2n; i++)
{
    f1();
    f2();

}

Code: Select all

f1()
{
   /*setup code left out*/
   magma_ssetmatrix ( m, m, a,  m, d_a ,  m );    // copy a -> d_a
   magmablas_slacpy ('A',m,m,d_a ,m,d_r ,m);     // copy d_a -> d_r
   magma_sgetrf_gpu( m, m, d_a, m, piv, &info);
   magma_sgetri_gpu(m,d_a,m,piv,dwork,ldwork,&info);

   /* transfer the result from gpu to cpu*/
   magma_sgetmatrix ( m, m, d_a , m, inv_result, m );
   

}

Code: Select all

 
f2()
{ 
        /* assuming that inv_result has global scope */
        stat = cublasSetVector (n, sizeof (float),inv_result ,1,d_x ,1); // cp inv_result->d_x
        cudaMemset(d_y,0,n*sizeof(float));
        stat=cublasSaxpy(handle,n,&epsilon,d_x,1,d_y,1);
        stat = cublasGetVector (n, sizeof ( float ),d_y ,1,matrix ,1);

}

[mgates3]

First, it is rarely advisable to invert a matrix. Solving a system is both faster and more accurate. Use gesv (which does getrf + getrs) to solve a system Ax=b, rather than inverting A and computing x = A^{-1} b.

I'm confused by two things.
First
- In f1, inv_result appears to be a M x M matrix, e.g., in magma_sgetmatrix.
- In f2, inv_result appears to be an N vector, e.g., in cublasSetVector.
So is it a matrix or a vector?

Second, at least in the code snippet you posted here, the inv_result is never used on the CPU, so why do you transfer it back to the CPU? Why not just leave it on the GPU?

Also, d_r is never used here.
-mark