The Portland Group | Support

PGI User Forum

Memberlist

Profile

CUDA Fortran : device variable in module
Goto page Previous 1, 2

PGI User Forum Forum Index -> Accelerator Programming

View previous topic :: View next topic

Author

Message

Tuan

Joined: 11 Jun 2009
Posts: 226

Posted: Thu Feb 18, 2010 12:53 pm Post subject:

goblinsqueen wrote:

Hi,
I receive the same warning than RTLEE when I compile the following code

Code:

module cudamod

use cudafor

real(kind=8), dimension(:), allocatable, device :: vec_dev

contains

attributes(device) subroutine sub1(i)

use cudafor

implicit none

integer, value :: i
real(kind=8), dimension(32) :: vec0
integer :: tid

tid = threadidx%x
vec0(tid) = vec_dev(i)

end subroutine

attributes(global) subroutine kernel_test(N,nblocks,nthreads)

use cudafor

implicit none

integer, value :: N, nthreads, nblocks
integer :: i, idx

idx = (blockidx%x-1)*blockdim%x + threadidx%x
do i=idx,N,nthreads*nblocks
call sub1(i)
end do

end subroutine

end module

program test

use cudafor
use cudamod

integer, parameter :: N = 1000
real(kind=8), dimension(N) :: vector
integer :: nblocks, nthreads

vector = 1.D0
allocate(vec_dev(N))
vec_dev = vector(1:N)
nblocks = 30
nthreads = 256
call kernel_test<<<nblocks,nthreads>>>(N,nblocks,nthreads)
deallocate(vec_dev)

end program

From the PGI CUDA Fortran documentation and from Mat's answer, I was assuming that it is not allowed for the kernel to access device data declared in a module, but that it is possible for any device subprogram in that module and by any host program that uses the module.

Am I wrong?

Thank you in advance for any explanation!

I experience the same issue. Don't access the device data from module.

Tuan.

mkcolg

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

Posted: Thu Feb 18, 2010 9:46 pm Post subject:

Hi Goblinqueen, Currently only fixed size device arrays are allowed in a Module. However, we are actively working on adding this (it's our most requested feature). It should be available in 10.6 or possibly sooner. - Mat

goblinsqueen

Joined: 04 Feb 2010
Posts: 14

Posted: Fri Feb 19, 2010 2:03 am Post subject:

Thank you, it would be very important!

Display posts from previous:

	PGI User Forum Forum Index -> Accelerator Programming	All times are GMT - 7 Hours Goto page Previous 1, 2
Page 2 of 2

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