libgomp/testsuite/libgomp.c/target-29.c

   1 #include <omp.h>
   2 #include <stdlib.h>
   3
   4 struct S { char p[64]; int a; int b[2]; long c[4]; int *d; char q[64]; };
   5
   6 __attribute__((noinline, noclone)) void
   7 foo (struct S s)
   8 {
   9   int d = omp_get_default_device ();
  10   int id = omp_get_initial_device ();
  11   int sep = 1;
  12
  13   if (d < 0 || d >= omp_get_num_devices ())
  14     d = id;
  15
  16   int err;
  17   #pragma omp target map(tofrom: s.a, s.b, s.c[1:2], s.d[-2:3]) map(to: sep) map(from: err)
  18   {
  19     err = s.a != 11 || s.b[0] != 12 || s.b[1] != 13;
  20     err |= s.c[1] != 15 || s.c[2] != 16 || s.d[-2] != 18 || s.d[-1] != 19 || s.d[0] != 20;
  21     s.a = 35; s.b[0] = 36; s.b[1] = 37;
  22     s.c[1] = 38; s.c[2] = 39; s.d[-2] = 40; s.d[-1] = 41; s.d[0] = 42;
  23     sep = 0;
  24   }
  25   if (err) abort ();
  26   err = s.a != 35 || s.b[0] != 36 || s.b[1] != 37;
  27   err |= s.c[1] != 38 || s.c[2] != 39 || s.d[-2] != 40 || s.d[-1] != 41 || s.d[0] != 42;
  28   if (err) abort ();
  29   s.a = 50; s.b[0] = 49; s.b[1] = 48;
  30   s.c[1] = 47; s.c[2] = 46; s.d[-2] = 45; s.d[-1] = 44; s.d[0] = 43;
  31   if (sep
  32       && (omp_target_is_present (&s.a, d)
  33           || omp_target_is_present (s.b, d)
  34           || omp_target_is_present (&s.c[1], d)
  35           || omp_target_is_present (s.d, d)
  36           || omp_target_is_present (&s.d[-2], d)))
  37     abort ();
  38   #pragma omp target data map(alloc: s.a, s.b, s.c[1:2], s.d[-2:3])
  39   {
  40     if (!omp_target_is_present (&s.a, d)
  41         || !omp_target_is_present (s.b, d)
  42         || !omp_target_is_present (&s.c[1], d)
  43         || !omp_target_is_present (s.d, d)
  44         || !omp_target_is_present (&s.d[-2], d))
  45       abort ();
  46     #pragma omp target update to(s.a, s.b, s.c[1:2], s.d[-2:3])
  47     #pragma omp target map(alloc: s.a, s.b, s.c[1:2], s.d[-2:3]) map(from: err)
  48     {
  49       err = s.a != 50 || s.b[0] != 49 || s.b[1] != 48;
  50       err |= s.c[1] != 47 || s.c[2] != 46 || s.d[-2] != 45 || s.d[-1] != 44 || s.d[0] != 43;
  51       s.a = 17; s.b[0] = 18; s.b[1] = 19;
  52       s.c[1] = 20; s.c[2] = 21; s.d[-2] = 22; s.d[-1] = 23; s.d[0] = 24;
  53     }
  54     #pragma omp target update from(s.a, s.b, s.c[1:2], s.d[-2:3])
  55   }
  56   if (sep
  57       && (omp_target_is_present (&s.a, d)
  58           || omp_target_is_present (s.b, d)
  59           || omp_target_is_present (&s.c[1], d)
  60           || omp_target_is_present (s.d, d)
  61           || omp_target_is_present (&s.d[-2], d)))
  62     abort ();
  63   if (err) abort ();
  64   err = s.a != 17 || s.b[0] != 18 || s.b[1] != 19;
  65   err |= s.c[1] != 20 || s.c[2] != 21 || s.d[-2] != 22 || s.d[-1] != 23 || s.d[0] != 24;
  66   if (err) abort ();
  67   s.a = 33; s.b[0] = 34; s.b[1] = 35;
  68   s.c[1] = 36; s.c[2] = 37; s.d[-2] = 38; s.d[-1] = 39; s.d[0] = 40;
  69   #pragma omp target enter data map(alloc: s.a, s.b, s.c[1:2], s.d[-2:3])
  70   if (!omp_target_is_present (&s.a, d)
  71       || !omp_target_is_present (s.b, d)
  72       || !omp_target_is_present (&s.c[1], d)
  73       || !omp_target_is_present (s.d, d)
  74       || !omp_target_is_present (&s.d[-2], d))
  75     abort ();
  76   #pragma omp target enter data map(always, to: s.a, s.b, s.c[1:2], s.d[-2:3])
  77   #pragma omp target map(alloc: s.a, s.b, s.c[1:2], s.d[-2:3]) map(from: err)
  78   {
  79     err = s.a != 33 || s.b[0] != 34 || s.b[1] != 35;
  80     err |= s.c[1] != 36 || s.c[2] != 37 || s.d[-2] != 38 || s.d[-1] != 39 || s.d[0] != 40;
  81     s.a = 49; s.b[0] = 48; s.b[1] = 47;
  82     s.c[1] = 46; s.c[2] = 45; s.d[-2] = 44; s.d[-1] = 43; s.d[0] = 42;
  83   }
  84   #pragma omp target exit data map(always, from: s.a, s.b, s.c[1:2], s.d[-2:3])
  85   if (!omp_target_is_present (&s.a, d)
  86       || !omp_target_is_present (s.b, d)
  87       || !omp_target_is_present (&s.c[1], d)
  88       || !omp_target_is_present (s.d, d)
  89       || !omp_target_is_present (&s.d[-2], d))
  90     abort ();
  91   #pragma omp target exit data map(release: s.a, s.b, s.c[1:2], s.d[-2:3])
  92   if (sep
  93       && (omp_target_is_present (&s.a, d)
  94           || omp_target_is_present (s.b, d)
  95           || omp_target_is_present (&s.c[1], d)
  96           || omp_target_is_present (s.d, d)
  97           || omp_target_is_present (&s.d[-2], d)))
  98     abort ();
  99   if (err) abort ();
 100   err = s.a != 49 || s.b[0] != 48 || s.b[1] != 47;
 101   err |= s.c[1] != 46 || s.c[2] != 45 || s.d[-2] != 44 || s.d[-1] != 43 || s.d[0] != 42;
 102   if (err) abort ();
 103 }
 104
 105 int
 106 main ()
 107 {
 108   int d[3] = { 18, 19, 20 };
 109   struct S s = { {}, 11, { 12, 13 }, { 14, 15, 16, 17 }, d + 2, {} };
 110   foo (s);
 111   return 0;
 112 }