88 lines · c
1// --------------------------------------------------2// Check extends before3// --------------------------------------------------4 5// RUN: %libomptarget-compile-generic \6// RUN: -DEXTENDS=BEFORE7// RUN: %libomptarget-run-fail-generic 2>&1 \8// RUN: | %fcheck-generic9 10// --------------------------------------------------11// Check extends after12// --------------------------------------------------13 14// RUN: %libomptarget-compile-generic \15// RUN: -DEXTENDS=AFTER16// RUN: %libomptarget-run-fail-generic 2>&1 \17// RUN: | %fcheck-generic18 19// END.20 21#include <stdio.h>22 23#define BEFORE 024#define AFTER 125 26#define SIZE 10027 28#if EXTENDS == BEFORE29#define SMALL_BEG (SIZE - 2)30#define SMALL_END SIZE31#define LARGE_BEG 032#define LARGE_END SIZE33#elif EXTENDS == AFTER34#define SMALL_BEG 035#define SMALL_END 236#define LARGE_BEG 037#define LARGE_END SIZE38#else39#error EXTENDS undefined40#endif41 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_SIZE48#define LARGE \49 LARGE_BEG: \50 LARGE_SIZE51 52int main() {53 int arr[SIZE];54 55 // CHECK: addr=0x[[#%x,SMALL_ADDR:]], size=[[#%u,SMALL_BYTES:]]56 fprintf(stderr, "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, "addr=%p, size=%ld\n", &arr[LARGE_BEG],61 LARGE_SIZE * sizeof arr[0]);62 63 // CHECK-NOT: omptarget64#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 present71 fprintf(stderr, "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 mandatory77#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 present84 fprintf(stderr, "arr is present\n");85 86 return 0;87}88