Skip to content

Commit 9aca391

Browse files
amitamd7amitamd7
committed
non-contiguous_update_to_tests
1 parent 97732dd commit 9aca391

File tree

6 files changed

+268
-0
lines changed

6 files changed

+268
-0
lines changed

‎offload/test/offloading/strided_multiple_update.c‎ renamed to ‎offload/test/offloading/strided_multiple_update_from.c‎

File renamed without changes.

‎offload/test/offloading/strided_multiple_update_to.c‎

Lines changed: 123 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,123 @@
1+
// This test checks that #pragma omp target update to(data1[0:3:4],
2+
// data2[0:2:5]) correctly updates disjoint strided sections of multiple arrays
3+
// from the host to the device.
4+
5+
// RUN: %libomptarget-compile-run-and-check-generic
6+
#include <omp.h>
7+
#include <stdio.h>
8+
9+
int main() {
10+
int len = 12;
11+
double data1[len], data2[len];
12+
13+
// Initialize host arrays
14+
for (int i = 0; i < len; i++) {
15+
data1[i] = i;
16+
data2[i] = i * 10;
17+
}
18+
19+
printf("original host array values:\n");
20+
printf("data1:\n");
21+
for (int i = 0; i < len; i++)
22+
printf("%.1f\n", data1[i]);
23+
printf("data2:\n");
24+
for (int i = 0; i < len; i++)
25+
printf("%.1f\n", data2[i]);
26+
27+
// CHECK: original host array values:
28+
// CHECK-NEXT: data1:
29+
// CHECK-NEXT: 0.0
30+
// CHECK-NEXT: 1.0
31+
// CHECK-NEXT: 2.0
32+
// CHECK-NEXT: 3.0
33+
// CHECK-NEXT: 4.0
34+
// CHECK-NEXT: 5.0
35+
// CHECK-NEXT: 6.0
36+
// CHECK-NEXT: 7.0
37+
// CHECK-NEXT: 8.0
38+
// CHECK-NEXT: 9.0
39+
// CHECK-NEXT: 10.0
40+
// CHECK-NEXT: 11.0
41+
// CHECK-NEXT: data2:
42+
// CHECK-NEXT: 0.0
43+
// CHECK-NEXT: 10.0
44+
// CHECK-NEXT: 20.0
45+
// CHECK-NEXT: 30.0
46+
// CHECK-NEXT: 40.0
47+
// CHECK-NEXT: 50.0
48+
// CHECK-NEXT: 60.0
49+
// CHECK-NEXT: 70.0
50+
// CHECK-NEXT: 80.0
51+
// CHECK-NEXT: 90.0
52+
// CHECK-NEXT: 100.0
53+
// CHECK-NEXT: 110.0
54+
55+
#pragma omp target data map(tofrom : data1[0 : len], data2[0 : len])
56+
{
57+
// Initialize device arrays to 20
58+
#pragma omp target
59+
{
60+
for (int i = 0; i < len; i++) {
61+
data1[i] = 20.0;
62+
data2[i] = 20.0;
63+
}
64+
}
65+
66+
// Modify host arrays for strided elements
67+
data1[0] = 10.0;
68+
data1[4] = 10.0;
69+
data1[8] = 10.0;
70+
data2[0] = 10.0;
71+
data2[5] = 10.0;
72+
73+
// data1[0:3:4] // indices 0,4,8
74+
// data2[0:2:5] // indices 0,5
75+
#pragma omp target update to(data1[0 : 3 : 4], data2[0 : 2 : 5])
76+
77+
// Verify on device by adding 5
78+
#pragma omp target
79+
{
80+
for (int i = 0; i < len; i++)
81+
data1[i] += 5.0;
82+
for (int i = 0; i < len; i++)
83+
data2[i] += 5.0;
84+
}
85+
}
86+
87+
printf("device array values after update to:\n");
88+
printf("data1:\n");
89+
for (int i = 0; i < len; i++)
90+
printf("%.1f\n", data1[i]);
91+
printf("data2:\n");
92+
for (int i = 0; i < len; i++)
93+
printf("%.1f\n", data2[i]);
94+
95+
// CHECK: device array values after update to:
96+
// CHECK-NEXT: data1:
97+
// CHECK-NEXT: 15.0
98+
// CHECK-NEXT: 25.0
99+
// CHECK-NEXT: 25.0
100+
// CHECK-NEXT: 25.0
101+
// CHECK-NEXT: 15.0
102+
// CHECK-NEXT: 25.0
103+
// CHECK-NEXT: 25.0
104+
// CHECK-NEXT: 25.0
105+
// CHECK-NEXT: 15.0
106+
// CHECK-NEXT: 25.0
107+
// CHECK-NEXT: 25.0
108+
// CHECK-NEXT: 25.0
109+
// CHECK-NEXT: data2:
110+
// CHECK-NEXT: 15.0
111+
// CHECK-NEXT: 25.0
112+
// CHECK-NEXT: 25.0
113+
// CHECK-NEXT: 25.0
114+
// CHECK-NEXT: 25.0
115+
// CHECK-NEXT: 15.0
116+
// CHECK-NEXT: 25.0
117+
// CHECK-NEXT: 25.0
118+
// CHECK-NEXT: 25.0
119+
// CHECK-NEXT: 25.0
120+
// CHECK-NEXT: 25.0
121+
// CHECK-NEXT: 25.0
122+
}
123+

‎offload/test/offloading/strided_partial_update.c‎ renamed to ‎offload/test/offloading/strided_partial_update_from.c‎

File renamed without changes.

‎offload/test/offloading/strided_partial_update_to.c‎

Lines changed: 73 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,73 @@
1+
// This test checks that #pragma omp target update to(data[0:4:3]) correctly
2+
// updates every third element (stride 3) from the host to the device, partially
3+
// across the array
4+
5+
// RUN: %libomptarget-compile-run-and-check-generic
6+
#include <omp.h>
7+
#include <stdio.h>
8+
9+
int main() {
10+
int len = 11;
11+
double data[len];
12+
13+
// Initialize on host
14+
for (int i = 0; i < len; i++)
15+
data[i] = i;
16+
17+
// Initial values
18+
printf("original host array values:\n");
19+
for (int i = 0; i < len; i++)
20+
printf("%f\n", data[i]);
21+
printf("\n");
22+
23+
// CHECK: 0.000000
24+
// CHECK: 1.000000
25+
// CHECK: 2.000000
26+
// CHECK: 3.000000
27+
// CHECK: 4.000000
28+
// CHECK: 5.000000
29+
// CHECK: 6.000000
30+
// CHECK: 7.000000
31+
// CHECK: 8.000000
32+
// CHECK: 9.000000
33+
// CHECK: 10.000000
34+
35+
#pragma omp target data map(tofrom : data[0 : len])
36+
{
37+
// Initialize device array to 20
38+
#pragma omp target
39+
for (int i = 0; i < len; i++)
40+
data[i] = 20.0;
41+
42+
// Modify host data for strided elements
43+
data[0] = 10.0;
44+
data[3] = 10.0;
45+
data[6] = 10.0;
46+
data[9] = 10.0;
47+
48+
#pragma omp target update to(data[0 : 4 : 3]) // indices 0,3,6,9
49+
50+
// Verify on device by adding 5
51+
#pragma omp target
52+
for (int i = 0; i < len; i++)
53+
data[i] += 5.0;
54+
}
55+
56+
printf("device array values after update to:\n");
57+
for (int i = 0; i < len; i++)
58+
printf("%f\n", data[i]);
59+
printf("\n");
60+
61+
// CHECK: 15.000000
62+
// CHECK: 25.000000
63+
// CHECK: 25.000000
64+
// CHECK: 15.000000
65+
// CHECK: 25.000000
66+
// CHECK: 25.000000
67+
// CHECK: 15.000000
68+
// CHECK: 25.000000
69+
// CHECK: 25.000000
70+
// CHECK: 15.000000
71+
// CHECK: 25.000000
72+
}
73+

‎offload/test/offloading/strided_update.c‎ renamed to ‎offload/test/offloading/strided_update_from.c‎

File renamed without changes.

‎offload/test/offloading/strided_update_to.c‎

Lines changed: 72 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,72 @@
1+
// This test checks that "update to" clause in OpenMP is supported when the
2+
// elements are updated in a non-contiguous manner. This test checks that
3+
// #pragma omp target update to(data[0:4:2]) correctly updates only every
4+
// other element (stride 2) from the host to the device
5+
6+
// RUN: %libomptarget-compile-run-and-check-generic
7+
#include <omp.h>
8+
#include <stdio.h>
9+
10+
int main() {
11+
int len = 8;
12+
double data[len];
13+
14+
// Initialize on host
15+
for (int i = 0; i < len; i++) {
16+
data[i] = i;
17+
}
18+
19+
// Initial values
20+
printf("original host array values:\n");
21+
for (int i = 0; i < len; i++)
22+
printf("%f\n", data[i]);
23+
printf("\n");
24+
25+
#pragma omp target data map(tofrom : len, data[0 : len])
26+
{
27+
// Initialize device to 20
28+
#pragma omp target
29+
for (int i = 0; i < len; i++) {
30+
data[i] = 20.0;
31+
}
32+
33+
// Modify host for strided elements
34+
data[0] = 10.0;
35+
data[2] = 10.0;
36+
data[4] = 10.0;
37+
data[6] = 10.0;
38+
39+
#pragma omp target update to(data[0 : 4 : 2])
40+
41+
// Verify on device by adding 5
42+
#pragma omp target
43+
for (int i = 0; i < len; i++) {
44+
data[i] += 5.0;
45+
}
46+
}
47+
48+
// CHECK: 0.000000
49+
// CHECK: 1.000000
50+
// CHECK: 2.000000
51+
// CHECK: 3.000000
52+
// CHECK: 4.000000
53+
// CHECK: 5.000000
54+
// CHECK: 6.000000
55+
// CHECK: 7.000000
56+
57+
printf("device array values after update to:\n");
58+
for (int i = 0; i < len; i++)
59+
printf("%f\n", data[i]);
60+
printf("\n");
61+
62+
// CHECK: 15.000000
63+
// CHECK: 25.000000
64+
// CHECK: 15.000000
65+
// CHECK: 25.000000
66+
// CHECK: 15.000000
67+
// CHECK: 25.000000
68+
// CHECK: 15.000000
69+
// CHECK: 25.000000
70+
71+
return 0;
72+
}

0 commit comments

Comments
 (0)