PGI User Forum
 SearchSearch   MemberlistMemberlist     RegisterRegister   ProfileProfile    Log inLog in 

Free OpenACC Webinar

swap pointers in device memory

 
Post new topic   Reply to topic    PGI User Forum Forum Index -> Accelerator Programming
View previous topic :: View next topic  
Author Message
Tuan



Joined: 11 Jun 2009
Posts: 233

PostPosted: Wed Jan 05, 2011 12:13 pm    Post subject: swap pointers in device memory Reply with quote

In CUDA C, we can easily swap pointers to two different memory region, e.g. in finite-difference computation
Code:

begin loop
   // this kernel update output based on input
   mykernek<<< .... >>>( float* output, float* input)
   float* tmp =  output;
   output = input;
   input = tmp;

end loop

Is there something similar that we can do with CUDA Fortran, or Fortran language.
What i can think of is to unroll the loop
Code:

begin loop
! unroll 1
   ! this kernel update output based on input
   mykernek<<< .... >>>( output,  input)
 
  ! do other stuff

! unroll 2
   ! this kernel update output based on input
   mykernek<<< .... >>>( input,  output)
 
  ! do other stuff

end loop

Is there a better way?

Thanks,

Tuan
Back to top
View user's profile
mkcolg



Joined: 30 Jun 2004
Posts: 6218
Location: The Portland Group Inc.

PostPosted: Wed Jan 05, 2011 5:45 pm    Post subject: Reply with quote

Quote:
Is there a better way?
We'll there isn't an easy way to swap pointers. I guess you could create ISO C c_ptr types but it seems to be an "un-Fortran" way to do it.

Your solution should work, but you could instead of unrolling, use an if statement to toggle which way to call the kernel based on the loop index. If odd call "output, input", if even call "input, output".

- Mat
Back to top
View user's profile
Tuan



Joined: 11 Jun 2009
Posts: 233

PostPosted: Thu Jan 06, 2011 7:44 am    Post subject: Reply with quote

mkcolg wrote:
Quote:
Is there a better way?
We'll there isn't an easy way to swap pointers. I guess you could create ISO C c_ptr types but it seems to be an "un-Fortran" way to do it.

- Mat


Will C_PTR work for pointer to device global memory when I try to use in Fortran?

Thanks Mat.


Tuan
Back to top
View user's profile
Display posts from previous:   
Post new topic   Reply to topic    PGI User Forum Forum Index -> Accelerator Programming All times are GMT - 7 Hours
Page 1 of 1

 
Jump to:  
You cannot post new topics in this forum
You cannot reply to topics in this forum
You cannot edit your posts in this forum
You cannot delete your posts in this forum
You cannot vote in polls in this forum


Powered by phpBB © phpBB Group