PGI User Forum
 SearchSearch   MemberlistMemberlist     RegisterRegister   ProfileProfile    Log inLog in 

Free OpenACC Webinar

host_data directive issue with PGI13.1

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



Joined: 16 Feb 2011
Posts: 69

PostPosted: Wed Jan 30, 2013 3:47 am    Post subject: host_data directive issue with PGI13.1 Reply with quote

Hi,

I am testing pgi13.1 which seems to support the host_data directive (there are no more error at compile time).

I tried it with a small test code which uses the host_data directive (see below), but it seems to only work when the -g flag is used:
Code:

> nvcc -c simple.cu
> pgf90 -acc -ta=nvidia,cuda5.0 -L$CUDALIB -lcudart -lcuda simple.o -o test.exe test.f90
> ./test.exe
call to cuStreamSynchronize returned error 700: Launch failed array a=0x69d860 b=0x69db80 n=100
srun: error: castor26: task 0: Exited with exit code 1


with -g
Code:

> nvcc -c simple.cu
> pgf90 -g -acc -ta=nvidia,cuda5.0 -L$CUDALIB -lcudart -lcuda simple.o -o
> ./test.exe
array a=0x900100200 b=0x900100000 n=100
    3.000000        12.00000   
    3.000000        12.00000



Here is the code:
Code:

> cat simple.cu

extern "C" {

__global__  void simple_add(float* a, float* b, int n) {
  int i = threadIdx.x+ blockIdx.x * blockDim.x;
  if ( i < n ) {
      a[i] = a[i] + b[i];
  }
}

void my_cuda_(float* a, float* b, int* p_n) {
    int n = *p_n;

    printf( "array a=%p b=%p n=%d\n", a, b, n );

    cudaThreadSynchronize( );

    simple_add<<<(n+127)/128, 128>>>(a,  b, n);

    cudaThreadSynchronize( );

}

};

----------
> cat test.f90
subroutine test(a,b,n)
  integer n
  real a(n),b(n)

!dir$ inlinenever test

!$acc data present(a,b)

!$acc host_data use_device(a,b)
  call my_cuda(a,b,n)
!$acc end host_data

!$acc end data

end subroutine test


program driver
  integer n
  real a(100),b(100),act_a(100)

  n = 100
  do i=1,n
     b(i) = real(i,4)
  end do
  a = 1.0

  act_a = a + b + 1.0

!$acc data copy(a),copyin(b)

!$acc parallel
  do i=1,n
     a(i) = a(i) + 1.0
  end do
!$acc end parallel

  call test(a,b,n)

!$acc end data

  print *, act_a(1), act_a(10)
  print *, a(1),a(10)

end program driver


Am I doing something wrong ?

Thanks,
Xavier
Back to top
View user's profile
xlapillonne



Joined: 16 Feb 2011
Posts: 69

PostPosted: Mon Feb 25, 2013 7:05 am    Post subject: Reply with quote

Hi,

Just wanted to know if there are any news concerning the host data directive.

I've just tried again the example above with pgi13.2 and I am still getting the same runtime error when not using the -g flag.

On your OpenAcc feature table host_data is still not officialy supported, do you have any timeline concerning this feature ?

Best regards,

Xavier
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