swap pointers in device memory

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

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

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

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

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

Thanks Mat.


Tuan