[OpenMP] Fix libomptarget test mapping/ompx_hold/struct.c

[jdenny]

Summary

For me, the test fails for nvptx64 offload. The problem was
introduced by D146838, which landed as 747af2415519. It tries to copy
a string constant's address from device to host and then print the
string. This patch copies the contents of the string instead.

openmp/libomptarget/test/mapping/ompx_hold/struct.c

 // RUN: %libomptarget-compile-generic -fopenmp-extensions
 // RUN: %libomptarget-run-generic | %fcheck-generic -strict-whitespace
 
 #include <omp.h>
 #include <stdio.h>
 
 #pragma omp begin declare target
-char *N1, *N2;
+#define MAX_NAME_SIZE 100
+char N1[MAX_NAME_SIZE], N2[MAX_NAME_SIZE];
 int V1, V2;
-#pragma omp declare target
+void copy_name(char *dst, char *src) {
+  int i;
+  for (i = 0; i < MAX_NAME_SIZE - 1 && src[i]; ++i)
+    dst[i] = src[i];
+  dst[i] = 0;
+}
+#pragma omp end declare target
 
 #define CHECK_PRESENCE(Var1, Var2, Var3)                                       \
   printf("    presence of %s, %s, %s: %d, %d, %d\n", #Var1, #Var2, #Var3,      \
          omp_target_is_present(&(Var1), omp_get_default_device()),             \
          omp_target_is_present(&(Var2), omp_get_default_device()),             \
          omp_target_is_present(&(Var3), omp_get_default_device()))
 
 #define CHECK_VALUES_HELPER(N1, N2, Var1, Var2)                                \
   printf("    values of %s, %s: %d, %d\n", N1, N2, (Var1), (Var2))
 
 #define CHECK_VALUES_DELAYED(Var1, Var2)                                       \
-  N1 = #Var1;                                                                  \
-  N2 = #Var2;                                                                  \
+  copy_name(N1, #Var1);                                                        \
+  copy_name(N2, #Var2);                                                        \
   V1 = (Var1);                                                                 \
   V2 = (Var2);
 
-#define CHECK_DELAYED_VALUS()                                                  \
+#define CHECK_DELAYED_VALUES()                                                 \
   _Pragma("omp target update from(N1, N2, V1, V2)")                            \
       CHECK_VALUES_HELPER(N1, N2, V1, V2)
 
 #define CHECK_VALUES(Var1, Var2)                                               \
   CHECK_VALUES_HELPER(#Var1, #Var2, (Var1), (Var2))
 
 int main() {
   struct S {
#pragma omp target map(from : s.i, s.j)
     // CHECK-NEXT: values of s.i, s.j: 20, 30
     CHECK_PRESENCE(s, s.i, s.j);
     CHECK_VALUES(s.i, s.j);
 #pragma omp target map(to : s.i, s.j)
     { // No transfer here even though parent's DynRefCount=1.
       // CHECK-NEXT: values of s.i, s.j: 21, 31
       CHECK_VALUES_DELAYED(s.i, s.j);
     }
-    CHECK_DELAYED_VALUS();
+    CHECK_DELAYED_VALUES();
   }
   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
   // CHECK-NEXT: values of s.i, s.j: 21, 31
   CHECK_PRESENCE(s, s.i, s.j);
   CHECK_VALUES(s.i, s.j);
 
   // -----------------------------------------------------------------------
   // CHECK-LABEL: check:{{.*}}
#pragma omp target map(ompx_hold, from : s.i, s.j)
     // CHECK-NEXT: values of s.i, s.j: 20, 30
     CHECK_PRESENCE(s, s.i, s.j);
     CHECK_VALUES(s.i, s.j);
 #pragma omp target map(ompx_hold, to : s.i, s.j)
     { // No transfer here even though parent's HoldRefCount=1.
       // CHECK-NEXT: values of s.i, s.j: 21, 31
       CHECK_VALUES_DELAYED(s.i, s.j);
     }
-    CHECK_DELAYED_VALUS();
+    CHECK_DELAYED_VALUES();
   }
   // CHECK-NEXT: presence of s, s.i, s.j: 0, 0, 0
   // CHECK-NEXT: values of s.i, s.j: 21, 31
   CHECK_PRESENCE(s, s.i, s.j);
   CHECK_VALUES(s.i, s.j);
 
   // -----------------------------------------------------------------------
   // CHECK-LABEL: check:{{.*}}