offload/test/mapping/ompx_hold/target-data.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 #define CHECK_PRESENCE(Var1, Var2, Var3)                                       \
   8   printf("    presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3,      \
   9          omp_target_is_present(&Var1, omp_get_default_device()),               \
  10          omp_target_is_present(&Var2, omp_get_default_device()),               \
  11          omp_target_is_present(&Var3, omp_get_default_device()))
  12
  13 int main() {
  14   int m, r, d;
  15   // CHECK: presence of m, r, d: 0, 0, 0
  16   CHECK_PRESENCE(m, r, d);
  17
  18   // -----------------------------------------------------------------------
  19   // CHECK-NEXT: check:{{.*}}
  20   printf("check: dyn>0, hold=0, dec/reset dyn=0\n");
  21
  22   // CHECK-NEXT: structured{{.*}}
  23   printf("  structured dec of dyn\n");
  24 #pragma omp target data map(tofrom : m) map(alloc : r, d)
  25   {
  26     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  27     CHECK_PRESENCE(m, r, d);
  28 #pragma omp target data map(tofrom : m) map(alloc : r, d)
  29     {
  30       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  31       CHECK_PRESENCE(m, r, d);
  32     }
  33     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  34     CHECK_PRESENCE(m, r, d);
  35   }
  36   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  37   CHECK_PRESENCE(m, r, d);
  38
  39   // CHECK-NEXT: dynamic{{.*}}
  40   printf("  dynamic dec/reset of dyn\n");
  41 #pragma omp target data map(tofrom : m) map(alloc : r, d)
  42   {
  43     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  44     CHECK_PRESENCE(m, r, d);
  45 #pragma omp target data map(tofrom : m) map(alloc : r, d)
  46     {
  47       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  48       CHECK_PRESENCE(m, r, d);
  49 #pragma omp target exit data map(from : m) map(release : r)
  50       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  51       CHECK_PRESENCE(m, r, d);
  52 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
  53       // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  54       CHECK_PRESENCE(m, r, d);
  55     }
  56     // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  57     CHECK_PRESENCE(m, r, d);
  58 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
  59     // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  60     CHECK_PRESENCE(m, r, d);
  61   }
  62   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  63   CHECK_PRESENCE(m, r, d);
  64
  65   // -----------------------------------------------------------------------
  66   // CHECK: check:{{.*}}
  67   printf("check: dyn=0, hold>0, dec/reset dyn=0, dec hold=0\n");
  68
  69   // Structured dec of dyn would require dyn>0.
  70
  71   // CHECK-NEXT: dynamic{{.*}}
  72   printf("  dynamic dec/reset of dyn\n");
  73 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
  74   {
  75     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  76     CHECK_PRESENCE(m, r, d);
  77 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
  78     {
  79       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  80       CHECK_PRESENCE(m, r, d);
  81 #pragma omp target exit data map(from : m) map(release : r)
  82       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  83       CHECK_PRESENCE(m, r, d);
  84 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
  85       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  86       CHECK_PRESENCE(m, r, d);
  87     }
  88     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  89     CHECK_PRESENCE(m, r, d);
  90 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
  91     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
  92     CHECK_PRESENCE(m, r, d);
  93   }
  94   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
  95   CHECK_PRESENCE(m, r, d);
  96
  97   // -----------------------------------------------------------------------
  98   // CHECK: check:{{.*}}
  99   printf("check: dyn>0, hold>0, dec/reset dyn=0, dec hold=0\n");
 100
 101   // CHECK-NEXT: structured{{.*}}
 102   printf("  structured dec of dyn\n");
 103 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 104   {
 105     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 106     CHECK_PRESENCE(m, r, d);
 107 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 108     {
 109       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 110       CHECK_PRESENCE(m, r, d);
 111 #pragma omp target data map(tofrom : m) map(alloc : r, d)
 112       {
 113         // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 114         CHECK_PRESENCE(m, r, d);
 115 #pragma omp target data map(tofrom : m) map(alloc : r, d)
 116         {
 117           // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 118           CHECK_PRESENCE(m, r, d);
 119         }
 120         // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 121         CHECK_PRESENCE(m, r, d);
 122       }
 123       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 124       CHECK_PRESENCE(m, r, d);
 125     }
 126     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 127     CHECK_PRESENCE(m, r, d);
 128   }
 129   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
 130   CHECK_PRESENCE(m, r, d);
 131
 132   // CHECK-NEXT: dynamic{{.*}}
 133   printf("  dynamic dec/reset of dyn\n");
 134 #pragma omp target enter data map(to : m) map(alloc : r, d)
 135   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 136   CHECK_PRESENCE(m, r, d);
 137 #pragma omp target enter data map(to : m) map(alloc : r, d)
 138   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 139   CHECK_PRESENCE(m, r, d);
 140 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 141   {
 142     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 143     CHECK_PRESENCE(m, r, d);
 144 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 145     {
 146       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 147       CHECK_PRESENCE(m, r, d);
 148 #pragma omp target exit data map(from : m) map(release : r)
 149       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 150       CHECK_PRESENCE(m, r, d);
 151 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
 152       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 153       CHECK_PRESENCE(m, r, d);
 154     }
 155     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 156     CHECK_PRESENCE(m, r, d);
 157 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
 158     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 159     CHECK_PRESENCE(m, r, d);
 160   }
 161   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
 162   CHECK_PRESENCE(m, r, d);
 163
 164   // -----------------------------------------------------------------------
 165   // CHECK: check:{{.*}}
 166   printf("check: dyn>0, hold>0, dec hold=0, dec/reset dyn=0\n");
 167
 168   // CHECK-NEXT: structured{{.*}}
 169   printf("  structured dec of dyn\n");
 170 #pragma omp target data map(tofrom : m) map(alloc : r, d)
 171   {
 172     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 173     CHECK_PRESENCE(m, r, d);
 174 #pragma omp target data map(tofrom : m) map(alloc : r, d)
 175     {
 176       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 177       CHECK_PRESENCE(m, r, d);
 178 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 179       {
 180         // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 181         CHECK_PRESENCE(m, r, d);
 182 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 183         {
 184           // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 185           CHECK_PRESENCE(m, r, d);
 186         }
 187         // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 188         CHECK_PRESENCE(m, r, d);
 189       }
 190       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 191       CHECK_PRESENCE(m, r, d);
 192     }
 193     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 194     CHECK_PRESENCE(m, r, d);
 195   }
 196   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
 197   CHECK_PRESENCE(m, r, d);
 198
 199   // CHECK-NEXT: dynamic{{.*}}
 200   printf("  dynamic dec/reset of dyn\n");
 201 #pragma omp target enter data map(to : m) map(alloc : r, d)
 202   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 203   CHECK_PRESENCE(m, r, d);
 204 #pragma omp target enter data map(to : m) map(alloc : r, d)
 205   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 206   CHECK_PRESENCE(m, r, d);
 207 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 208   {
 209     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 210     CHECK_PRESENCE(m, r, d);
 211 #pragma omp target data map(ompx_hold, tofrom : m) map(ompx_hold, alloc : r, d)
 212     {
 213       // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 214       CHECK_PRESENCE(m, r, d);
 215     }
 216     // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 217     CHECK_PRESENCE(m, r, d);
 218   }
 219   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 220   CHECK_PRESENCE(m, r, d);
 221 #pragma omp target exit data map(from : m) map(release : r)
 222   // CHECK-NEXT: presence of m, r, d: 1, 1, 1
 223   CHECK_PRESENCE(m, r, d);
 224 #pragma omp target exit data map(from : m) map(release : r) map(delete : d)
 225   // CHECK-NEXT: presence of m, r, d: 0, 0, 0
 226   CHECK_PRESENCE(m, r, d);
 227
 228   return 0;
 229 }