1// RUN: %libomptarget-compile-run-and-check-generic
2
3// REQUIRES: libc
4
5#include <assert.h>
6#include <omp.h>
7#include <stdio.h>
8
9#pragma omp begin declare variant match(device = {kind(gpu)})
10// Extension provided by the 'libc' project.
11unsigned long long __llvm_omp_host_call(void *fn, void *args, size_t size);
12#pragma omp declare target to(__llvm_omp_host_call) device_type(nohost)
13#pragma omp end declare variant
14
15#pragma omp begin declare variant match(device = {kind(cpu)})
16// Dummy host implementation to make this work for all targets.
17unsigned long long __llvm_omp_host_call(void *fn, void *args, size_t size) {
18 return ((unsigned long long (*)(void *))fn)(args);
19}
20#pragma omp end declare variant
21
22long long foo(void *data) { return -1; }
23
24void *fn_ptr = NULL;
25#pragma omp declare target to(fn_ptr)
26
27int main() {
28 fn_ptr = (void *)&foo;
29#pragma omp target update to(fn_ptr)
30
31 for (int i = 0; i < 4; ++i) {
32#pragma omp target
33 {
34 long long res = __llvm_omp_host_call(fn: fn_ptr, NULL, size: 0);
35 assert(res == -1 && "RPC call failed\n");
36 }
37
38 for (int j = 0; j < 128; ++j) {
39#pragma omp target nowait
40 {
41 long long res = __llvm_omp_host_call(fn: fn_ptr, NULL, size: 0);
42 assert(res == -1 && "RPC call failed\n");
43 }
44 }
45#pragma omp taskwait
46
47#pragma omp target
48 {
49 long long res = __llvm_omp_host_call(fn: fn_ptr, NULL, size: 0);
50 assert(res == -1 && "RPC call failed\n");
51 }
52 }
53
54 // CHECK: PASS
55 puts(s: "PASS");
56}
57

source code of offload/test/libc/server.c