libgomp/testsuite/libgomp.oacc-c-c++-common/host_data-1.c

   1 /* { dg-do run { target openacc_nvidia_accel_selected } } */
   2 /* { dg-additional-options "-lcuda -lcublas -lcudart" } */
   3
   4 #include <stdlib.h>
   5 #include <openacc.h>
   6 #include <cuda.h>
   7 #include <cuda_runtime_api.h>
   8 #include <cublas_v2.h>
   9
  10 void
  11 saxpy_host (int n, float a, float *x, float *y)
  12 {
  13   int i;
  14
  15   for (i = 0; i < n; i++)
  16     y[i] = y[i] + a * x[i];
  17 }
  18
  19 #pragma acc routine
  20 void
  21 saxpy_target (int n, float a, float *x, float *y)
  22 {
  23   int i;
  24
  25   for (i = 0; i < n; i++)
  26     y[i] = y[i] + a * x[i];
  27 }
  28
  29 int
  30 main(int argc, char **argv)
  31 {
  32 #define N 8
  33   int i;
  34   float x_ref[N], y_ref[N];
  35   float x[N], y[N];
  36   cublasHandle_t h;
  37   float a = 2.0;
  38
  39   for (i = 0; i < N; i++)
  40     {
  41       x[i] = x_ref[i] = 4.0 + i;
  42       y[i] = y_ref[i] = 3.0;
  43     }
  44
  45   saxpy_host (N, a, x_ref, y_ref);
  46
  47   cublasCreate (&h);
  48
  49 #pragma acc data copyin (x[0:N]) copy (y[0:N])
  50   {
  51 #pragma acc host_data use_device (x, y)
  52     {
  53       cublasSaxpy (h, N, &a, x, 1, y, 1);
  54     }
  55   }
  56
  57   for (i = 0; i < N; i++)
  58     {
  59       if (y[i] != y_ref[i])
  60         abort ();
  61     }
  62
  63 #pragma acc data create (x[0:N]) copyout (y[0:N])
  64   {
  65 #pragma acc kernels
  66     for (i = 0; i < N; i++)
  67       y[i] = 3.0;
  68
  69 #pragma acc host_data use_device (x, y)
  70     {
  71       cublasSaxpy (h, N, &a, x, 1, y, 1);
  72     }
  73   }
  74
  75   cublasDestroy (h);
  76
  77   for (i = 0; i < N; i++)
  78     {
  79       if (y[i] != y_ref[i])
  80         abort ();
  81     }
  82
  83   for (i = 0; i < N; i++)
  84     y[i] = 3.0;
  85
  86   /* There's no need to use host_data here.  */
  87 #pragma acc data copyin (x[0:N]) copyin (a) copy (y[0:N])
  88   {
  89 #pragma acc parallel present (x[0:N]) pcopy (y[0:N]) present (a)
  90     saxpy_target (N, a, x, y);
  91   }
  92
  93   for (i = 0; i < N; i++)
  94     {
  95       if (y[i] != y_ref[i])
  96         abort ();
  97     }
  98
  99   return 0;
 100 }