PGI User Forum
 SearchSearch   MemberlistMemberlist     RegisterRegister   ProfileProfile    Log inLog in 

CUDA-x86.

[CUDA FORTRAN] Problem with SHARED attribute

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



Joined: 06 Jul 2011
Posts: 27

PostPosted: Tue Oct 04, 2011 7:01 am    Post subject: [CUDA FORTRAN] Problem with SHARED attribute Reply with quote

Hello,

I've written a simple 2D reduction code in order to fin the minimum value of a 16*16 matrix. I found a problem with the "shared" attribute: in particular, if I use a "device" array the code performs correctly, instead with the "shared" attribute it doesn't works (the result is displayed as a row of stars).

I've uploaded the CUDA FORTRAN project on MediaFire:

http://www.mediafire.com/?h1qoz5g8velrcte

Note: to perform the 2D reduction with the shared memory feature, just uncomment the call the the second kernel and comment the call to the first one (using device memory).


Thanks in advance,

Nicola
Back to top
View user's profile
mkcolg



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

PostPosted: Tue Oct 04, 2011 2:10 pm    Post subject: Reply with quote

Hi Nicola,

You're using dynamic shared memory (shared automatic arrays) but failed to specify the size of the memory in the kernel launch. Hence, your kernel is crashing and 'res' is returning garbage.

To fix, either change your shared arrays to be fixed size (i.e. change dim_mx_x and dim_mx_y to be parameters) or pass in the shared memory size. For example:
Code:
call simple_red_shrd  1,dimBlock_s_mx,((dim_mx_x*dim_mx_y*pr)+(dim_mx_x*2*pr))   (dim_mx_x,dim_mx_y,s_mx_dev,res)


Hope this helps,
Mat
Back to top
View user's profile
nicolaprandi



Joined: 06 Jul 2011
Posts: 27

PostPosted: Tue Oct 04, 2011 11:56 pm    Post subject: Reply with quote

mkcolg wrote:
(i.e. change dim_mx_x and dim_mx_y to be parameters)


Thanks for you help, I chose this way and fixed the problem. In order to help other people who may read this thread, what I needed to do was the following:

- Declare the 2D array's dimensions (dim_mx_x, dim_mx_y) as parameters in the variables module:
Code:
integer,parameter :: dim_mx_x = 16,dim_mx_y = 16

- Remove the dimensions from the kernel call:
Code:
call simple_red_shrd<<<1>>>(s_mx_dev,res)

- Inside the kernel, declare the dimensions with a "USE ONLY" statement:
Code:
use simple_variables,only: dim_mx_x,dim_mx_y



Nicola
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