offload/test/mapping/ompx_hold/struct.c

   1 // RUN: %libomptarget-compile-generic -fopenmp-extensions
   2 // RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
   3
   4 #include <omp.h>
   5 #include <stdio.h>
   6
   7 #pragma omp begin declare target
   8 #define MAX_NAME_SIZE 100
   9 char N1[MAX_NAME_SIZE], N2[MAX_NAME_SIZE];
  10 int V1, V2;
  11 void copy_name(char *dst, char *src) {
  12   int i;
  13   for (i = 0; i < MAX_NAME_SIZE - 1 && src[i]; ++i)
  14     dst[i] = src[i];
  15   dst[i] = 0;
  16 }
  17 #pragma omp end declare target
  18
  19 #define CHECK_PRESENCE(Var1, Var2, Var3)                                       \
  20   printf("    presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3,      \
  21          omp_target_is_present(&(Var1), omp_get_default_device()),             \
  22          omp_target_is_present(&(Var2), omp_get_default_device()),             \
  23          omp_target_is_present(&(Var3), omp_get_default_device()))
  24
  25 #define CHECK_VALUES_HELPER(N1, N2, Var1, Var2)                                \
  26   printf("    values of %s, %s: %d, %d\n", N1, N2, (Var1), (Var2))
  27
  28 #define CHECK_VALUES_DELAYED(Var1, Var2)                                       \
  29   copy_name(N1, #Var1);                                                        \
  30   copy_name(N2, #Var2);                                                        \
  31   V1 = (Var1);                                                                 \
  32   V2 = (Var2);
  33
  34 #define CHECK_DELAYED_VALUES()                                                 \
  35   _Pragma("omp target update from(N1, N2, V1, V2)")                            \
  36       CHECK_VALUES_HELPER(N1, N2, V1, V2)
  37
  38 #define CHECK_VALUES(Var1, Var2)                                               \
  39   CHECK_VALUES_HELPER(#Var1, #Var2, (Var1), (Var2))
  40
  41 int main() {
  42   struct S {
  43     int i;
  44     int j;
  45   } s;
  46   // CHECK: presence of s, s.i, s.j: 0, 0, 0
  47   CHECK_PRESENCE(s, s.i, s.j);
  48
  49   // =======================================================================
  50   // Check that ompx_hold keeps entire struct present.
  51
  52   // -----------------------------------------------------------------------
  53   // CHECK-LABEL: check:{{.*}}
  54   printf("check: ompx_hold only on first member\n");
  55   s.i = 20;
  56   s.j = 30;
  57 #pragma omp target data map(tofrom : s) map(ompx_hold, tofrom : s.i)           \
  58     map(tofrom : s.j)
  59   {
  60     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
  61     CHECK_PRESENCE(s, s.i, s.j);
  62 #pragma omp target map(tofrom : s)
  63     {
  64       s.i = 21;
  65       s.j = 31;
  66     }
  67 #pragma omp target exit data map(delete : s, s.i)
  68     // ompx_hold on s.i applies to all of s.
  69     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
  70     // CHECK-NEXT: values of s.i, s.j: 20, 30
  71     CHECK_PRESENCE(s, s.i, s.j);
  72     CHECK_VALUES(s.i, s.j);
  73   }
  74   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
  75   // CHECK-NEXT: values of s.i, s.j: 21, 31
  76   CHECK_PRESENCE(s, s.i, s.j);
  77   CHECK_VALUES(s.i, s.j);
  78
  79   // -----------------------------------------------------------------------
  80   // CHECK-LABEL: check:{{.*}}
  81   printf("check: ompx_hold only on last member\n");
  82   s.i = 20;
  83   s.j = 30;
  84 #pragma omp target data map(tofrom : s) map(tofrom : s.i)                      \
  85     map(ompx_hold, tofrom : s.j)
  86   {
  87     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
  88     CHECK_PRESENCE(s, s.i, s.j);
  89 #pragma omp target map(tofrom : s)
  90     {
  91       s.i = 21;
  92       s.j = 31;
  93     }
  94 #pragma omp target exit data map(delete : s, s.i)
  95     // ompx_hold on s.j applies to all of s.
  96     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
  97     // CHECK-NEXT: values of s.i, s.j: 20, 30
  98     CHECK_PRESENCE(s, s.i, s.j);
  99     CHECK_VALUES(s.i, s.j);
 100   }
 101   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
 102   // CHECK-NEXT: values of s.i, s.j: 21, 31
 103   CHECK_PRESENCE(s, s.i, s.j);
 104   CHECK_VALUES(s.i, s.j);
 105
 106   // -----------------------------------------------------------------------
 107   // CHECK-LABEL: check:{{.*}}
 108   printf("check: ompx_hold only on struct\n");
 109   s.i = 20;
 110   s.j = 30;
 111 #pragma omp target data map(ompx_hold, tofrom : s) map(tofrom : s.i)           \
 112     map(tofrom : s.j)
 113   {
 114     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 115     CHECK_PRESENCE(s, s.i, s.j);
 116 #pragma omp target map(tofrom : s)
 117     {
 118       s.i = 21;
 119       s.j = 31;
 120     }
 121 #pragma omp target exit data map(delete : s, s.i)
 122     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 123     // CHECK-NEXT: values of s.i, s.j: 20, 30
 124     CHECK_PRESENCE(s, s.i, s.j);
 125     CHECK_VALUES(s.i, s.j);
 126   }
 127   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
 128   // CHECK-NEXT: values of s.i, s.j: 21, 31
 129   CHECK_PRESENCE(s, s.i, s.j);
 130   CHECK_VALUES(s.i, s.j);
 131
 132   // =======================================================================
 133   // Check that transfer to/from host checks reference count correctly.
 134
 135   // -----------------------------------------------------------------------
 136   // CHECK-LABEL: check:{{.*}}
 137   printf("check: parent DynRefCount=1 is not sufficient for transfer\n");
 138   s.i = 20;
 139   s.j = 30;
 140 #pragma omp target data map(ompx_hold, tofrom : s)
 141 #pragma omp target data map(ompx_hold, tofrom : s)
 142   {
 143     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 144     CHECK_PRESENCE(s, s.i, s.j);
 145 #pragma omp target map(from : s.i, s.j)
 146     {
 147       s.i = 21;
 148       s.j = 31;
 149     } // No transfer here even though parent's DynRefCount=1.
 150     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 151     // CHECK-NEXT: values of s.i, s.j: 20, 30
 152     CHECK_PRESENCE(s, s.i, s.j);
 153     CHECK_VALUES(s.i, s.j);
 154 #pragma omp target map(to : s.i, s.j)
 155     { // No transfer here even though parent's DynRefCount=1.
 156       // CHECK-NEXT: values of s.i, s.j: 21, 31
 157       CHECK_VALUES_DELAYED(s.i, s.j);
 158     }
 159     CHECK_DELAYED_VALUES();
 160   }
 161   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
 162   // CHECK-NEXT: values of s.i, s.j: 21, 31
 163   CHECK_PRESENCE(s, s.i, s.j);
 164   CHECK_VALUES(s.i, s.j);
 165
 166   // -----------------------------------------------------------------------
 167   // CHECK-LABEL: check:{{.*}}
 168   printf("check: parent HoldRefCount=1 is not sufficient for transfer\n");
 169   s.i = 20;
 170   s.j = 30;
 171 #pragma omp target data map(tofrom : s)
 172 #pragma omp target data map(tofrom : s)
 173   {
 174     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 175     CHECK_PRESENCE(s, s.i, s.j);
 176 #pragma omp target map(ompx_hold, from : s.i, s.j)
 177     {
 178       s.i = 21;
 179       s.j = 31;
 180     } // No transfer here even though parent's HoldRefCount=1.
 181     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 182     // CHECK-NEXT: values of s.i, s.j: 20, 30
 183     CHECK_PRESENCE(s, s.i, s.j);
 184     CHECK_VALUES(s.i, s.j);
 185 #pragma omp target map(ompx_hold, to : s.i, s.j)
 186     { // No transfer here even though parent's HoldRefCount=1.
 187       // CHECK-NEXT: values of s.i, s.j: 21, 31
 188       CHECK_VALUES_DELAYED(s.i, s.j);
 189     }
 190     CHECK_DELAYED_VALUES();
 191   }
 192   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
 193   // CHECK-NEXT: values of s.i, s.j: 21, 31
 194   CHECK_PRESENCE(s, s.i, s.j);
 195   CHECK_VALUES(s.i, s.j);
 196
 197   // -----------------------------------------------------------------------
 198   // CHECK-LABEL: check:{{.*}}
 199   //
 200   // At the beginning of a region, if the parent's TotalRefCount=1, then the
 201   // transfer should happen.
 202   //
 203   // At the end of a region, it also must be true that the reference count being
 204   // decremented is the reference count that is 1.
 205   printf("check: parent TotalRefCount=1 is not sufficient for transfer\n");
 206   s.i = 20;
 207   s.j = 30;
 208 #pragma omp target data map(ompx_hold, tofrom : s)
 209   {
 210     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 211     CHECK_PRESENCE(s, s.i, s.j);
 212 #pragma omp target map(ompx_hold, tofrom : s.i, s.j)
 213     {
 214       s.i = 21;
 215       s.j = 31;
 216     }
 217 #pragma omp target exit data map(from : s.i, s.j)
 218     // No transfer here even though parent's TotalRefCount=1.
 219     // CHECK-NEXT: presence of s, s.i, s.j: 1, 1, 1
 220     // CHECK-NEXT: values of s.i, s.j: 20, 30
 221     CHECK_PRESENCE(s, s.i, s.j);
 222     CHECK_VALUES(s.i, s.j);
 223   }
 224   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
 225   // CHECK-NEXT: values of s.i, s.j: 21, 31
 226   CHECK_PRESENCE(s, s.i, s.j);
 227   CHECK_VALUES(s.i, s.j);
 228
 229   return 0;
 230 }