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

source code of offload/test/mapping/present/target_data_array_extension.c