cudaopenacccufft

How to use acc_set_cuda_stream(streamId, stream)?

I created a CUDA stream in this way:

integer(kind=cuda_stream_kind) :: stream1
istat = cudaStreamCreate(stream1)

to use it for the plan of a cufft:

err_dir = err_dir + cufftPlan2D(plan_dir1,NY,NY,CUFFT_D2Z)
err_dir = err_dir + cufftSetStream(plan_dir1,stream1)

In the routine that executes the cufft, I pass plan_dir1 and I have

subroutine new_fft_dir(z,plan)

  !$acc host_data use_device(z)
  ierr = ierr + cufftExecD2Z(plan,z,z)
  !$acc end host_data

  !$acc parallel loop collapse(2) present(z)
  do i=1,NXP2
  do j=1,NY
  z(i,j) = z(i,j)/NY**2
  enddo
  enddo
  !$acc end parallel loop

I would like to set an OpenACC stream equal to the CUDA stream stream1, but using :

integer(kind=cuda_stream_kind) :: stream1
istat = cudaStreamCreate(stream1)
integer :: stream
istat = cudaStreamCreate(stream1)
acc_set_cuda_stream(stream,stream1)

I get **NVFORTRAN-S-0034-Syntax error at or near end of line (main.f90: 48) **

My goal is to add the async clause to

  !$acc parallel loop collapse(2) present(z) async(stream)
  do i=1,NXP2
  do j=1,NY
  z(i,j) = z(i,j)/NY**2
  enddo
  enddo
  !$acc end parallel loop

to have this loop and the fft on the same CUDA stream.

Could the problem be that I use integer(kind=cuda_stream_kind) intead of cudaStream_t stream?

Solution

  • "acc_set_cuda_stream" is a subroutine so you do need to add "call " before it. Also, variables need to be declared before executable code, hence "integer :: stream" needs to be moved up a line.

       use cudafor
   use openacc
   integer(kind=cuda_stream_kind) :: stream1
   integer :: stream
   istat = cudaStreamCreate(stream1)
   call acc_set_cuda_stream(stream,stream1)