[Offload][Tests] Non-contiguous_update_to_tests

clang/test/OpenMP/target_update_strided_messages.c → clang/test/OpenMP/target_update_strided_messages_from.c

@@ -1,8 +1,6 @@
 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
 
-void foo(void) {}
-
 int main(int argc, char **argv) {
   int len = 8;
   double data[len];
@@ -11,9 +9,6 @@ int main(int argc, char **argv) {
   #pragma omp target update from(data[0:4:2]) // OK
   {}
 
-  #pragma omp target update to(data[0:len/2:2]) // OK
-  {}
-
   #pragma omp target update from(data[1:3:2]) // OK
   {}
 
@@ -35,4 +30,5 @@ int main(int argc, char **argv) {
   {}
 
   return 0;
-}
+}
+

clang/test/OpenMP/target_update_strided_messages_to.c

@@ -0,0 +1,34 @@
+// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
+// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
+
+int main(int argc, char **argv) {
+  int len = 8;
+  double data[len];
+
+  // Valid strided array sections
+  #pragma omp target update to(data[0:4:2]) // OK
+  {}
+
+  #pragma omp target update to(data[1:3:2]) // OK
+  {}
+
+  // Missing stride (default = 1)
+  #pragma omp target update to(data[0:4]) // OK
+  {}
+
+  // Invalid stride expressions
+  #pragma omp target update to(data[0:4:0]) // expected-error {{section stride is evaluated to a non-positive value 0}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+
+  #pragma omp target update to(data[0:4:-1]) // expected-error {{section stride is evaluated to a non-positive value -1}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+
+  // Missing colon 
+  #pragma omp target update to(data[0:4 2]) // expected-error {{expected ']'}} expected-note {{to match this '['}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+  {}
+
+  // Too many colons
+  #pragma omp target update to(data[0:4:2:1]) // expected-error {{expected ']'}} expected-note {{to match this '['}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+  {}
+
+  return 0;
+}
+

clang/test/OpenMP/target_update_strided_multiple_messages.c → clang/test/OpenMP/target_update_strided_multiple_messages_from.c

@@ -1,46 +1,24 @@
 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
 
-void foo(void) {}
-
-typedef struct {
-  int len;
-  double data[12];
-} S;
-
 int main(int argc, char **argv) {
   int len = 12;
   double data1[len], data2[len];
-  S s;
 
   // Valid multiple strided array sections
   #pragma omp target update from(data1[0:4:2], data2[0:2:5]) // OK
   {}
 
-  #pragma omp target update to(data1[1:2:3], data2[2:3:2]) // OK
-  {}
-
   // Mixed strided and regular array sections
   #pragma omp target update from(data1[0:len], data2[0:4:2]) // OK
   {}
 
-  // Struct member arrays with strides
-  #pragma omp target update from(s.data[0:4:2]) // OK
-  {}
-
-  #pragma omp target update from(s.data[0:s.len/2:2]) // OK
-  {}
-
   // Invalid stride in one of multiple sections
   #pragma omp target update from(data1[0:3:4], data2[0:2:0]) // expected-error {{section stride is evaluated to a non-positive value 0}}
 
-  // Complex expressions in multiple arrays
-  int stride1 = 2, stride2 = 3;
-  #pragma omp target update from(data1[0:len/2:stride1], data2[1:len/3:stride2]) // OK
-  {}
-
   // Missing colon 
   #pragma omp target update from(data1[0:4:2], data2[0:3 4]) // expected-error {{expected ']'}} expected-note {{to match this '['}}
 
   return 0;
-}
+}
+

clang/test/OpenMP/target_update_strided_multiple_messages_to.c

@@ -0,0 +1,26 @@
+// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
+// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
+
+int main(int argc, char **argv) {
+  int len = 12;
+  double data1[len], data2[len];
+
+  // Valid multiple strided array sections
+  #pragma omp target update to(data1[0:4:2], data2[0:2:5]) // OK
+  {}
+
+  // Mixed strided and regular array sections
+  #pragma omp target update to(data1[0:len], data2[0:4:2]) // OK
+  {}
+
+  // Invalid stride in one of multiple sections
+  #pragma omp target update to(data1[0:3:4], data2[0:2:0]) // expected-error {{section stride is evaluated to a non-positive value 0}}
+  {}
+
+  // Missing colon 
+  #pragma omp target update to(data1[0:4:2], data2[0:3 4]) // expected-error {{expected ']'}} expected-note {{to match this '['}}
+  {}
+
+  return 0;
+}
+

clang/test/OpenMP/target_update_strided_partial_messages.c → clang/test/OpenMP/target_update_strided_partial_messages_from.c

@@ -1,7 +1,6 @@
 // RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
 // RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
-
-void foo(void) {}
+// expected-no-diagnostics
 
 int main(int argc, char **argv) {
   int len = 11;
@@ -17,16 +16,11 @@ int main(int argc, char **argv) {
 
   // Valid: complex expressions
   int offset = 1;
-  int count = 3;
-  int stride = 2;
-  #pragma omp target update from(data[offset:count:stride]) // OK
-  {}
-
+
   // Invalid stride expressions
   #pragma omp target update from(data[0:4:offset-1]) // OK if offset > 1
   {}
 
-  #pragma omp target update from(data[0:count:0]) // expected-error {{section stride is evaluated to a non-positive value 0}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-
   return 0;
-}
+}
+

clang/test/OpenMP/target_update_strided_partial_messages_to.c

@@ -0,0 +1,26 @@
+// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wuninitialized
+// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wuninitialized
+// expected-no-diagnostics
+
+int main(int argc, char **argv) {
+  int len = 11;
+  double data[len];
+
+  // Valid partial strided updates
+  #pragma omp target update to(data[0:4:3]) // OK
+  {}
+
+  // Stride larger than length
+  #pragma omp target update to(data[0:2:10]) // OK
+  {}
+
+  // Valid: complex expressions
+  int offset = 1;
+
+  // Potentially invalid stride expressions depending on runtime values
+  #pragma omp target update to(data[0:4:offset-1]) // OK if offset > 1
+  {}
+
+  return 0;
+}
+

offload/test/offloading/strided_multiple_update_to.c

@@ -0,0 +1,122 @@
+// This test checks that #pragma omp target update to(data1[0:3:4],
+// data2[0:2:5]) correctly updates disjoint strided sections of multiple arrays
+// from the host to the device.
+
+// RUN: %libomptarget-compile-run-and-check-generic
+#include <omp.h>
+#include <stdio.h>
+
+int main() {
+  int len = 12;
+  double data1[len], data2[len];
+
+  // Initialize host arrays
+  for (int i = 0; i < len; i++) {
+    data1[i] = i;
+    data2[i] = i * 10;
+  }
+
+  printf("original host array values:\n");
+  printf("data1:\n");
+  for (int i = 0; i < len; i++)
+    printf("%.1f\n", data1[i]);
+  printf("data2:\n");
+  for (int i = 0; i < len; i++)
+    printf("%.1f\n", data2[i]);
+
+  // CHECK: original host array values:
+  // CHECK-NEXT: data1:
+  // CHECK-NEXT: 0.0
+  // CHECK-NEXT: 1.0
+  // CHECK-NEXT: 2.0
+  // CHECK-NEXT: 3.0
+  // CHECK-NEXT: 4.0
+  // CHECK-NEXT: 5.0
+  // CHECK-NEXT: 6.0
+  // CHECK-NEXT: 7.0
+  // CHECK-NEXT: 8.0
+  // CHECK-NEXT: 9.0
+  // CHECK-NEXT: 10.0
+  // CHECK-NEXT: 11.0
+  // CHECK-NEXT: data2:
+  // CHECK-NEXT: 0.0
+  // CHECK-NEXT: 10.0
+  // CHECK-NEXT: 20.0
+  // CHECK-NEXT: 30.0
+  // CHECK-NEXT: 40.0
+  // CHECK-NEXT: 50.0
+  // CHECK-NEXT: 60.0
+  // CHECK-NEXT: 70.0
+  // CHECK-NEXT: 80.0
+  // CHECK-NEXT: 90.0
+  // CHECK-NEXT: 100.0
+  // CHECK-NEXT: 110.0
+
+#pragma omp target data map(tofrom : data1[0 : len], data2[0 : len])
+  {
+    // Initialize device arrays to 20
+#pragma omp target
+    {
+      for (int i = 0; i < len; i++) {
+        data1[i] = 20.0;
+        data2[i] = 20.0;
+      }
+    }
+
+    // Modify host arrays for strided elements
+    data1[0] = 10.0;
+    data1[4] = 10.0;
+    data1[8] = 10.0;
+    data2[0] = 10.0;
+    data2[5] = 10.0;
+
+    // data1[0:3:4]  // indices 0,4,8
+    // data2[0:2:5]  // indices 0,5
+#pragma omp target update to(data1[0 : 3 : 4], data2[0 : 2 : 5])
+
+    // Verify on device by adding 5
+#pragma omp target
+    {
+      for (int i = 0; i < len; i++)
+        data1[i] += 5.0;
+      for (int i = 0; i < len; i++)
+        data2[i] += 5.0;
+    }
+  }
+
+  printf("device array values after update to:\n");
+  printf("data1:\n");
+  for (int i = 0; i < len; i++)
+    printf("%.1f\n", data1[i]);
+  printf("data2:\n");
+  for (int i = 0; i < len; i++)
+    printf("%.1f\n", data2[i]);
+
+  // CHECK: device array values after update to:
+  // CHECK-NEXT: data1:
+  // CHECK-NEXT: 15.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 15.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 15.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: data2:
+  // CHECK-NEXT: 15.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 15.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+  // CHECK-NEXT: 25.0
+}

offload/test/offloading/strided_partial_update_to.c

@@ -0,0 +1,72 @@
+// This test checks that #pragma omp target update to(data[0:4:3]) correctly
+// updates every third element (stride 3) from the host to the device, partially
+// across the array
+
+// RUN: %libomptarget-compile-run-and-check-generic
+#include <omp.h>
+#include <stdio.h>
+
+int main() {
+  int len = 11;
+  double data[len];
+
+  // Initialize on host
+  for (int i = 0; i < len; i++)
+    data[i] = i;
+
+  // Initial values
+  printf("original host array values:\n");
+  for (int i = 0; i < len; i++)
+    printf("%f\n", data[i]);
+  printf("\n");
+
+  // CHECK: 0.000000
+  // CHECK: 1.000000
+  // CHECK: 2.000000
+  // CHECK: 3.000000
+  // CHECK: 4.000000
+  // CHECK: 5.000000
+  // CHECK: 6.000000
+  // CHECK: 7.000000
+  // CHECK: 8.000000
+  // CHECK: 9.000000
+  // CHECK: 10.000000
+
+#pragma omp target data map(tofrom : data[0 : len])
+  {
+    // Initialize device array to 20
+#pragma omp target
+    for (int i = 0; i < len; i++)
+      data[i] = 20.0;
+
+    // Modify host data for strided elements
+    data[0] = 10.0;
+    data[3] = 10.0;
+    data[6] = 10.0;
+    data[9] = 10.0;
+
+#pragma omp target update to(data[0 : 4 : 3]) // indices 0,3,6,9
+
+    // Verify on device by adding 5
+#pragma omp target
+    for (int i = 0; i < len; i++)
+      data[i] += 5.0;
+  }
+
+  printf("device array values after update to:\n");
+  for (int i = 0; i < len; i++)
+    printf("%f\n", data[i]);
+  printf("\n");
+
+  // CHECK: 15.000000
+  // CHECK: 25.000000
+  // CHECK: 25.000000
+  // CHECK: 15.000000
+  // CHECK: 25.000000
+  // CHECK: 25.000000
+  // CHECK: 15.000000
+  // CHECK: 25.000000
+  // CHECK: 25.000000
+  // CHECK: 15.000000
+  // CHECK: 25.000000
+}