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
main()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