1 | // RUN: %libomptarget-compile-run-and-check-generic |
2 | // REQUIRES: ompt |
3 | // REQUIRES: gpu |
4 | |
5 | /* |
6 | * Example OpenMP program that registers EMI callbacks |
7 | */ |
8 | |
9 | #include <assert.h> |
10 | #include <omp.h> |
11 | #include <stdio.h> |
12 | |
13 | #include "callbacks.h" |
14 | #include "register_emi.h" |
15 | |
16 | int main() { |
17 | int N = 100000; |
18 | |
19 | int a[N]; |
20 | int b[N]; |
21 | |
22 | int i; |
23 | |
24 | for (i = 0; i < N; i++) |
25 | a[i] = 0; |
26 | |
27 | for (i = 0; i < N; i++) |
28 | b[i] = i; |
29 | |
30 | #pragma omp target parallel for |
31 | { |
32 | for (int j = 0; j < N; j++) |
33 | a[j] = b[j]; |
34 | } |
35 | |
36 | #pragma omp target teams distribute parallel for |
37 | { |
38 | for (int j = 0; j < N; j++) |
39 | a[j] = b[j]; |
40 | } |
41 | |
42 | int rc = 0; |
43 | for (i = 0; i < N; i++) |
44 | if (a[i] != b[i]) { |
45 | rc++; |
46 | printf(format: "Wrong value: a[%d]=%d\n" , i, a[i]); |
47 | } |
48 | |
49 | if (!rc) |
50 | printf(format: "Success\n" ); |
51 | |
52 | return rc; |
53 | } |
54 | |
55 | /// CHECK: Callback Init: |
56 | /// CHECK: Callback Load: |
57 | /// CHECK: Callback Target EMI: kind=1 endpoint=1 |
58 | /// CHECK-NOT: code=(nil) |
59 | /// CHECK: code=[[CODE1:.*]] |
60 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=1 |
61 | /// CHECK: code=[[CODE1]] |
62 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=1 |
63 | /// CHECK-NOT: dest=(nil) |
64 | /// CHECK: code=[[CODE1]] |
65 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=2 |
66 | /// CHECK: code=[[CODE1]] |
67 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=2 |
68 | /// CHECK: code=[[CODE1]] |
69 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=1 |
70 | /// CHECK: code=[[CODE1]] |
71 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=1 |
72 | /// CHECK-NOT: dest=(nil) |
73 | /// CHECK: code=[[CODE1]] |
74 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=2 |
75 | /// CHECK: code=[[CODE1]] |
76 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=2 |
77 | /// CHECK: code=[[CODE1]] |
78 | /// CHECK: Callback Submit EMI: endpoint=1 req_num_teams=1 |
79 | /// CHECK: Callback Submit EMI: endpoint=2 req_num_teams=1 |
80 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=3 |
81 | /// CHECK: code=[[CODE1]] |
82 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=3 |
83 | /// CHECK: code=[[CODE1]] |
84 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=3 |
85 | /// CHECK: code=[[CODE1]] |
86 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=3 |
87 | /// CHECK: code=[[CODE1]] |
88 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=4 |
89 | /// CHECK: code=[[CODE1]] |
90 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=4 |
91 | /// CHECK: code=[[CODE1]] |
92 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=4 |
93 | /// CHECK: code=[[CODE1]] |
94 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=4 |
95 | /// CHECK: code=[[CODE1]] |
96 | /// CHECK: Callback Target EMI: kind=1 endpoint=2 |
97 | /// CHECK: code=[[CODE1]] |
98 | |
99 | /// CHECK: Callback Target EMI: kind=1 endpoint=1 |
100 | /// CHECK-NOT: code=(nil) |
101 | /// CHECK: code=[[CODE2:.*]] |
102 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=1 |
103 | /// CHECK: code=[[CODE2]] |
104 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=1 |
105 | /// CHECK-NOT: dest=(nil) |
106 | /// CHECK: code=[[CODE2]] |
107 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=2 |
108 | /// CHECK: code=[[CODE2]] |
109 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=2 |
110 | /// CHECK: code=[[CODE2]] |
111 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=1 |
112 | /// CHECK: code=[[CODE2]] |
113 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=1 |
114 | /// CHECK-NOT: dest=(nil) |
115 | /// CHECK: code=[[CODE2]] |
116 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=2 |
117 | /// CHECK: code=[[CODE2]] |
118 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=2 |
119 | /// CHECK: code=[[CODE2]] |
120 | /// CHECK: Callback Submit EMI: endpoint=1 req_num_teams=0 |
121 | /// CHECK: Callback Submit EMI: endpoint=2 req_num_teams=0 |
122 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=3 |
123 | /// CHECK: code=[[CODE2]] |
124 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=3 |
125 | /// CHECK: code=[[CODE2]] |
126 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=3 |
127 | /// CHECK: code=[[CODE2]] |
128 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=3 |
129 | /// CHECK: code=[[CODE2]] |
130 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=4 |
131 | /// CHECK: code=[[CODE2]] |
132 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=4 |
133 | /// CHECK: code=[[CODE2]] |
134 | /// CHECK: Callback DataOp EMI: endpoint=1 optype=4 |
135 | /// CHECK: code=[[CODE2]] |
136 | /// CHECK: Callback DataOp EMI: endpoint=2 optype=4 |
137 | /// CHECK: code=[[CODE2]] |
138 | /// CHECK: Callback Target EMI: kind=1 endpoint=2 |
139 | /// CHECK: code=[[CODE2]] |
140 | /// CHECK: Callback Fini: |
141 | |