annotate libgomp/testsuite/libgomp.c-c++-common/target-1.c @ 158:494b0b89df80 default tip

...
author Shinji KONO <kono@ie.u-ryukyu.ac.jp>
date Mon, 25 May 2020 18:13:55 +0900
parents 84e7813d76e9
children
Ignore whitespace changes - Everywhere: Within whitespace: At end of lines:
rev   line source
131
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
1 extern
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
2 #ifdef __cplusplus
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
3 "C"
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
4 #endif
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
5 void abort (void);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
6
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
7 void
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
8 fn1 (double *x, double *y, int z)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
9 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
10 int i;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
11 for (i = 0; i < z; i++)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
12 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
13 x[i] = i & 31;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
14 y[i] = (i & 63) - 30;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
15 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
16 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
17
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
18 #pragma omp declare target
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
19 int tgtv = 6;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
20 int
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
21 tgt (void)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
22 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
23 #pragma omp atomic update
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
24 tgtv++;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
25 return 0;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
26 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
27 #pragma omp end declare target
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
28
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
29 double
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
30 fn2 (int x, int y, int z)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
31 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
32 double b[1024], c[1024], s = 0;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
33 int i, j;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
34 fn1 (b, c, x);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
35 #pragma omp target data map(to: b)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
36 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
37 #pragma omp target map(tofrom: c, s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
38 #pragma omp teams num_teams(y) thread_limit(z) reduction(+:s) firstprivate(x)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
39 #pragma omp distribute dist_schedule(static, 4) collapse(1)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
40 for (j=0; j < x; j += y)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
41 #pragma omp parallel for reduction(+:s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
42 for (i = j; i < j + y; i++)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
43 tgt (), s += b[i] * c[i];
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
44 #pragma omp target update from(b, tgtv)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
45 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
46 return s;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
47 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
48
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
49 double
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
50 fn3 (int x)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
51 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
52 double b[1024], c[1024], s = 0;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
53 int i;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
54 fn1 (b, c, x);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
55 #pragma omp target map(to: b, c) map(tofrom:s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
56 #pragma omp parallel for reduction(+:s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
57 for (i = 0; i < x; i++)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
58 tgt (), s += b[i] * c[i];
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
59 return s;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
60 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
61
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
62 double
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
63 fn4 (int x, double *p)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
64 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
65 double b[1024], c[1024], d[1024], s = 0;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
66 int i;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
67 fn1 (b, c, x);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
68 fn1 (d + x, p + x, x);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
69 #pragma omp target map(to: b, c[0:x], d[x:x]) map(to:p[x:64 + (x & 31)]) \
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
70 map(tofrom: s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
71 #pragma omp parallel for reduction(+:s)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
72 for (i = 0; i < x; i++)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
73 s += b[i] * c[i] + d[x + i] + p[x + i];
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
74 return s;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
75 }
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
76
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
77 int
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
78 main ()
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
79 {
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
80 double a = fn2 (128, 4, 6);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
81 int b = tgtv;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
82 double c = fn3 (61);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
83 #pragma omp target update from(tgtv)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
84 int d = tgtv;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
85 double e[1024];
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
86 double f = fn4 (64, e);
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
87 if (a != 13888.0 || b != 6 + 128 || c != 4062.0 || d != 6 + 128 + 61
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
88 || f != 8032.0)
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
89 abort ();
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
90 return 0;
84e7813d76e9 gcc-8.2
mir3636
parents:
diff changeset
91 }