1 // --------------------------------------------------
2 // Check extends before
3 // --------------------------------------------------
4 
5 // RUN: %libomptarget-compile-generic \
6 // RUN:   -fopenmp-version=51 -DEXTENDS=BEFORE
7 // RUN: %libomptarget-run-fail-generic 2>&1 \
8 // RUN: | %fcheck-generic
9 
10 // --------------------------------------------------
11 // Check extends after
12 // --------------------------------------------------
13 
14 // RUN: %libomptarget-compile-generic \
15 // RUN:   -fopenmp-version=51 -DEXTENDS=AFTER
16 // RUN: %libomptarget-run-fail-generic 2>&1 \
17 // RUN: | %fcheck-generic
18 
19 
20 // END.
21 
22 #include <stdio.h>
23 
24 #define BEFORE 0
25 #define AFTER  1
26 
27 #define SIZE 100
28 
29 #if EXTENDS == BEFORE
30 # define SMALL_BEG (SIZE-2)
31 # define SMALL_END SIZE
32 # define LARGE_BEG 0
33 # define LARGE_END SIZE
34 #elif EXTENDS == AFTER
35 # define SMALL_BEG 0
36 # define SMALL_END 2
37 # define LARGE_BEG 0
38 # define LARGE_END SIZE
39 #else
40 # error EXTENDS undefined
41 #endif
42 
43 #define SMALL_SIZE (SMALL_END-SMALL_BEG)
44 #define LARGE_SIZE (LARGE_END-LARGE_BEG)
45 
46 #define SMALL SMALL_BEG:SMALL_SIZE
47 #define LARGE LARGE_BEG:LARGE_SIZE
48 
49 int main() {
50   int arr[SIZE];
51 
52   // CHECK: addr=0x[[#%x,SMALL_ADDR:]], size=[[#%u,SMALL_BYTES:]]
53   fprintf(stderr, "addr=%p, size=%ld\n", &arr[SMALL_BEG],
54           SMALL_SIZE * sizeof arr[0]);
55 
56   // CHECK: addr=0x[[#%x,LARGE_ADDR:]], size=[[#%u,LARGE_BYTES:]]
57   fprintf(stderr, "addr=%p, size=%ld\n", &arr[LARGE_BEG],
58           LARGE_SIZE * sizeof arr[0]);
59 
60   // CHECK-NOT: Libomptarget
61 #pragma omp target data map(alloc: arr[LARGE])
62   {
63 #pragma omp target data map(present, tofrom: arr[SMALL])
64     ;
65   }
66 
67   // CHECK: arr is present
68   fprintf(stderr, "arr is present\n");
69 
70   // CHECK: Libomptarget message: explicit extension not allowed: host address specified is 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes), but device allocation maps to host at 0x{{0*}}[[#SMALL_ADDR]] ([[#SMALL_BYTES]] bytes)
71   // CHECK: Libomptarget message: device mapping required by 'present' map type modifier does not exist for host address 0x{{0*}}[[#LARGE_ADDR]] ([[#LARGE_BYTES]] bytes)
72   // CHECK: Libomptarget error: Call to getTargetPointer returned null pointer ('present' map type modifier).
73   // CHECK: Libomptarget fatal error 1: failure of target construct while offloading is mandatory
74 #pragma omp target data map(alloc: arr[SMALL])
75   {
76 #pragma omp target data map(present, tofrom: arr[LARGE])
77     ;
78   }
79 
80   // CHECK-NOT: arr is present
81   fprintf(stderr, "arr is present\n");
82 
83   return 0;
84 }
85