230 lines · c
1// RUN: %libomptarget-compile-generic -fopenmp-extensions2// RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace3 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 13int main() {14 int m, r, d;15 // CHECK: presence of m, r, d: 0, 0, 016 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, 127 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, 131 CHECK_PRESENCE(m, r, d);32 }33 // CHECK-NEXT: presence of m, r, d: 1, 1, 134 CHECK_PRESENCE(m, r, d);35 }36 // CHECK-NEXT: presence of m, r, d: 0, 0, 037 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, 144 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, 148 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, 151 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, 054 CHECK_PRESENCE(m, r, d);55 }56 // CHECK-NEXT: presence of m, r, d: 0, 0, 057 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, 060 CHECK_PRESENCE(m, r, d);61 }62 // CHECK-NEXT: presence of m, r, d: 0, 0, 063 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, 176 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, 180 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, 183 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, 186 CHECK_PRESENCE(m, r, d);87 }88 // CHECK-NEXT: presence of m, r, d: 1, 1, 189 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, 192 CHECK_PRESENCE(m, r, d);93 }94 // CHECK-NEXT: presence of m, r, d: 0, 0, 095 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, 1106 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, 1110 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, 1114 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, 1118 CHECK_PRESENCE(m, r, d);119 }120 // CHECK-NEXT: presence of m, r, d: 1, 1, 1121 CHECK_PRESENCE(m, r, d);122 }123 // CHECK-NEXT: presence of m, r, d: 1, 1, 1124 CHECK_PRESENCE(m, r, d);125 }126 // CHECK-NEXT: presence of m, r, d: 1, 1, 1127 CHECK_PRESENCE(m, r, d);128 }129 // CHECK-NEXT: presence of m, r, d: 0, 0, 0130 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, 1136 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, 1139 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, 1143 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, 1147 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, 1150 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, 1153 CHECK_PRESENCE(m, r, d);154 }155 // CHECK-NEXT: presence of m, r, d: 1, 1, 1156 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, 1159 CHECK_PRESENCE(m, r, d);160 }161 // CHECK-NEXT: presence of m, r, d: 0, 0, 0162 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, 1173 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, 1177 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, 1181 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, 1185 CHECK_PRESENCE(m, r, d);186 }187 // CHECK-NEXT: presence of m, r, d: 1, 1, 1188 CHECK_PRESENCE(m, r, d);189 }190 // CHECK-NEXT: presence of m, r, d: 1, 1, 1191 CHECK_PRESENCE(m, r, d);192 }193 // CHECK-NEXT: presence of m, r, d: 1, 1, 1194 CHECK_PRESENCE(m, r, d);195 }196 // CHECK-NEXT: presence of m, r, d: 0, 0, 0197 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, 1203 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, 1206 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, 1210 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, 1214 CHECK_PRESENCE(m, r, d);215 }216 // CHECK-NEXT: presence of m, r, d: 1, 1, 1217 CHECK_PRESENCE(m, r, d);218 }219 // CHECK-NEXT: presence of m, r, d: 1, 1, 1220 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, 1223 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, 0226 CHECK_PRESENCE(m, r, d);227 228 return 0;229}230