1 // RUN: %libomptarget-compile-run-and-check-generic
2 // REQUIRES: ompt
3 // REQUIRES: gpu
4
5 /*
6 * Example OpenMP program that registers non-EMI callbacks
7 */
8
9 #include <omp.h>
10 #include <stdio.h>
11
12 #include "callbacks.h"
13 #include "register_non_emi.h"
14
main()15 int main() {
16 int N = 100000;
17
18 int a[N];
19 int b[N];
20
21 int i;
22
23 for (i = 0; i < N; i++)
24 a[i] = 0;
25
26 for (i = 0; i < N; i++)
27 b[i] = i;
28
29 #pragma omp target parallel for
30 {
31 for (int j = 0; j < N; j++)
32 a[j] = b[j];
33 }
34
35 #pragma omp target teams distribute parallel for
36 {
37 for (int j = 0; j < N; j++)
38 a[j] = b[j];
39 }
40
41 int rc = 0;
42 for (i = 0; i < N; i++)
43 if (a[i] != b[i]) {
44 rc++;
45 printf("Wrong value: a[%d]=%d\n", i, a[i]);
46 }
47
48 if (!rc)
49 printf("Success\n");
50
51 return rc;
52 }
53
54 // clang-format off
55 /// CHECK: Callback Init:
56 /// CHECK: Callback Load:
57 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1 device_num=[[DEVICE_NUM:[0-9]+]]
58 /// CHECK-NOT: code=(nil)
59 /// CHECK: code=[[CODE1:.*]]
60 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
61 /// CHECK: code=[[CODE1]]
62 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
63 /// CHECK: code=[[CODE1]]
64 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
65 /// CHECK: code=[[CODE1]]
66 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
67 /// CHECK: code=[[CODE1]]
68 /// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=1
69 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
70 /// CHECK: code=[[CODE1]]
71 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
72 /// CHECK: code=[[CODE1]]
73 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
74 /// CHECK: code=[[CODE1]]
75 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
76 /// CHECK: code=[[CODE1]]
77 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 device_num=[[DEVICE_NUM]] code=[[CODE1]]
78
79 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=1
80 /// device_num=[[DEVICE_NUM]]
81 /// CHECK-NOT: code=(nil)
82 /// CHECK: code=[[CODE2:.*]]
83 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
84 /// CHECK: code=[[CODE2]]
85 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
86 /// CHECK: code=[[CODE2]]
87 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=1
88 /// CHECK: code=[[CODE2]]
89 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=2
90 /// CHECK: code=[[CODE2]]
91 /// CHECK: Callback Submit: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] req_num_teams=0
92 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
93 /// CHECK: code=[[CODE2]]
94 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=3
95 /// CHECK: code=[[CODE2]]
96 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
97 /// CHECK: code=[[CODE2]]
98 /// CHECK: Callback DataOp: target_id=[[TARGET_ID:[0-9]+]] host_op_id=[[HOST_OP_ID:[0-9]+]] optype=4
99 /// CHECK: code=[[CODE2]]
100 /// CHECK: Callback Target: target_id=[[TARGET_ID:[0-9]+]] kind=1 endpoint=2 device_num=[[DEVICE_NUM]] code=[[CODE2]]
101 /// CHECK: Callback Fini:
102