cchen updated this revision to Diff 243950.
cchen added a comment.
1. Refactor based on feedback and use the visitor class to integrate the
analysis
in checkMapClauseExpressionBase.
2. Handle more lvalue cases
Repository:
rG LLVM Github Monorepo
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D72811/new/
https://reviews.llvm.org/D72811
Files:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/lib/CodeGen/CGOpenMPRuntime.cpp
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/target_map_messages.cpp
clang/test/OpenMP/target_teams_map_messages.cpp
clang/test/OpenMP/target_update_codegen.cpp
clang/test/OpenMP/target_update_from_messages.cpp
clang/test/OpenMP/target_update_to_messages.cpp
Index: clang/test/OpenMP/target_update_to_messages.cpp
===================================================================
--- clang/test/OpenMP/target_update_to_messages.cpp
+++ clang/test/OpenMP/target_update_to_messages.cpp
@@ -1,6 +1,12 @@
-// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp -fopenmp-version=50 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-simd -fopenmp-version=50 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
void foo() {
}
@@ -59,6 +65,29 @@
double *p;
unsigned bfa : 4;
};
+struct S8 {
+ int *ptr;
+ int a;
+ struct S7* S;
+ void foo() {
+#pragma omp target update to(*this) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+ {}
+#pragma omp target update to(*(&(*this))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(this->ptr)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(&(*(this->S->i+this->S->p)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(&(*(this->S->i+this->S->s6[0].pp)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(&(*(a+this->ptr)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(&(*(*(this->ptr)+a+this->ptr)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+ {}
+ }
+};
+
+struct Base {
+ virtual ~Base() {}
+};
+struct Derived: Base {
+ virtual void name() {}
+};
S3 h;
#pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}}
@@ -89,12 +118,12 @@
#pragma omp target update to(x)
#pragma omp target update to(t[:I])
#pragma omp target update to(T) // expected-error {{'T' does not refer to a value}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update to(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update to(I) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update to(S2::S2s)
#pragma omp target update to(S2::S2sc)
#pragma omp target update to(to)
#pragma omp target update to(y x) // expected-error {{expected ',' or ')' in 'to' clause}}
-#pragma omp target update to(argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update to(argc > 0 ? x : y) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update to(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update to(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
#pragma omp target update to(ba)
@@ -106,7 +135,7 @@
#pragma omp target update to(x, a[:2]) // expected-error {{subscripted value is not an array or pointer}}
#pragma omp target update to(x, c[:]) // expected-error {{subscripted value is not an array or pointer}}
-#pragma omp target update to(x, (m+1)[2]) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update to(x, (m+1)[2]) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update to(s7.i, s7.a[:3])
#pragma omp target update to(s7.s6[1].aa[0:5])
#pragma omp target update to(x, s7.s6[:5].aa[6]) // expected-error {{OpenMP array section is not allowed here}}
@@ -118,6 +147,16 @@
{
#pragma omp target update to(s7.x)
}
+
+ const int &rci = i;
+ Base *b = new Base;
+ int *pd = const_cast<int*>(&d);
+#pragma omp target update to(const_cast<int&>(rci))
+#pragma omp target update to(const_cast<int*>(pd))
+#pragma omp target update to(static_cast<double>(i))
+#pragma omp target update to(dynamic_cast<Derived*>(b))
+#pragma omp target update to(reinterpret_cast<int*>(b))
+ {}
return 0;
}
@@ -135,6 +174,8 @@
const int (&l)[5] = da;
S7 s7;
int *m;
+ int **BB, *offset;
+ int ***BBB;
#pragma omp target update to // expected-error {{expected '(' after 'to'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update to( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
@@ -147,7 +188,7 @@
#pragma omp target update to(S2::S2sc)
#pragma omp target update to(to)
#pragma omp target update to(y x) // expected-error {{expected ',' or ')' in 'to' clause}}
-#pragma omp target update to(argc > 0 ? x : y) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(argc > 0 ? x : y) // le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update to(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update to(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
#pragma omp target update to(ba)
@@ -159,7 +200,7 @@
#pragma omp target update to(x, a[:2]) // expected-error {{subscripted value is not an array or pointer}}
#pragma omp target update to(x, c[:]) // expected-error {{subscripted value is not an array or pointer}}
-#pragma omp target update to(x, (m+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update to(x, (m+1)[2]) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update to(s7.i, s7.a[:3])
#pragma omp target update to(s7.s6[1].aa[0:5])
#pragma omp target update to(x, s7.s6[:5].aa[6]) // expected-error {{OpenMP array section is not allowed here}}
@@ -172,6 +213,32 @@
#pragma omp target update to(s7.x)
}
+#pragma omp target update to(*(&y)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(&(*(&y)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(BB+*offset)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(BB+y)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(m+*offset)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(*offset+BB)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(y+BB)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(*offset+m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(*offset+BB+*m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(**(*(*(&offset))+BB+*m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*(x+*(y+*(**BB+BBB)+s7.i)+static_cast<int>(*(s7.s6[0].pp)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+ {}
+
+ const int &rci = i;
+ Base *b = new Base;
+ int *pd = const_cast<int*>(&d);
+#pragma omp target update to(const_cast<int&>(rci))
+#pragma omp target update to(const_cast<int*>(pd))
+#pragma omp target update to(static_cast<double>(i))
+#pragma omp target update to(dynamic_cast<Derived*>(b))
+#pragma omp target update to(reinterpret_cast<int*>(b))
+ {}
+
+#pragma omp target update to(static_cast<double>(i + 1)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}} le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
+ {}
+
return tmain<int, 3>(argc)+tmain<from, 4>(argc); // expected-note {{in instantiation of function template specialization 'tmain<int, 3>' requested here}} expected-note {{in instantiation of function template specialization 'tmain<int, 4>' requested here}}
}
Index: clang/test/OpenMP/target_update_from_messages.cpp
===================================================================
--- clang/test/OpenMP/target_update_from_messages.cpp
+++ clang/test/OpenMP/target_update_from_messages.cpp
@@ -1,6 +1,9 @@
-// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp -fopenmp-version=50 -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
void foo() {
}
@@ -59,6 +62,30 @@
double *p;
unsigned bfa : 4;
};
+struct S8 {
+ int *ptr;
+ int a;
+ struct S7* S;
+ void foo() {
+#pragma omp target update from(*this) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+ {}
+#pragma omp target update from(*(&(*this))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(this->ptr)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(&(*(this->S->i+this->S->p)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(&(*(this->S->i+this->S->s6[0].pp)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(&(*(a+this->ptr)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(&(*(*(this->ptr)+a+this->ptr)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+ {}
+ }
+};
+
+struct Base {
+ virtual ~Base() {}
+};
+struct Derived: Base {
+ virtual void name() {}
+};
+
S3 h;
#pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}}
@@ -89,12 +116,13 @@
#pragma omp target update from(x)
#pragma omp target update from(t[:I])
#pragma omp target update from(T) // expected-error {{'T' does not refer to a value}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
-#pragma omp target update from(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update from(I) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update from(S2::S2s)
#pragma omp target update from(S2::S2sc)
#pragma omp target update from(from)
#pragma omp target update from(y x) // expected-error {{expected ',' or ')' in 'from' clause}}
-#pragma omp target update from(argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update from(argc > 0 ? x : y) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+
#pragma omp target update from(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update from(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
#pragma omp target update from(ba)
@@ -106,7 +134,7 @@
#pragma omp target update from(x, a[:2]) // expected-error {{subscripted value is not an array or pointer}}
#pragma omp target update from(x, c[:]) // expected-error {{subscripted value is not an array or pointer}}
-#pragma omp target update from(x, (m+1)[2]) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update from(x, (m+1)[2]) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update from(s7.i, s7.a[:3])
#pragma omp target update from(s7.s6[1].aa[0:5])
#pragma omp target update from(x, s7.s6[:5].aa[6]) // expected-error {{OpenMP array section is not allowed here}}
@@ -119,6 +147,16 @@
#pragma omp target update from(s7.x)
}
+ const int &rci = i;
+ Base *b = new Base;
+ int *pd = const_cast<int*>(&d);
+#pragma omp target update to(const_cast<int&>(rci))
+#pragma omp target update to(const_cast<int*>(pd))
+#pragma omp target update to(static_cast<double>(i))
+#pragma omp target update to(dynamic_cast<Derived*>(b))
+#pragma omp target update to(reinterpret_cast<int*>(b))
+ {}
+
return 0;
}
@@ -136,6 +174,8 @@
const int (&l)[5] = da;
int *m;
S7 s7;
+ int **BB, *offset;
+ int ***BBB;
#pragma omp target update from // expected-error {{expected '(' after 'from'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update from( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
@@ -148,7 +188,7 @@
#pragma omp target update from(S2::S2sc)
#pragma omp target update from(from)
#pragma omp target update from(y x) // expected-error {{expected ',' or ')' in 'from' clause}}
-#pragma omp target update from(argc > 0 ? x : y) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(argc > 0 ? x : y) // le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update from(S1) // expected-error {{'S1' does not refer to a value}}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target update from(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
#pragma omp target update from(ba)
@@ -160,7 +200,7 @@
#pragma omp target update from(x, a[:2]) // expected-error {{subscripted value is not an array or pointer}}
#pragma omp target update from(x, c[:]) // expected-error {{subscripted value is not an array or pointer}}
-#pragma omp target update from(x, (m+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update from(x, (m+1)[2]) // // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target update from(s7.i, s7.a[:3])
#pragma omp target update from(s7.s6[1].aa[0:5])
#pragma omp target update from(x, s7.s6[:5].aa[6]) // expected-error {{OpenMP array section is not allowed here}}
@@ -168,11 +208,35 @@
#pragma omp target update from(s7.p[:10])
#pragma omp target update from(x, s7.bfa) // expected-error {{bit fields cannot be used to specify storage in a 'from' clause}}
#pragma omp target update from(x, s7.p[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
+#pragma omp target update from(*(&y)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(&(*(&y)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(BB+*offset)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(BB+y)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(m+*offset)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(*offset+BB)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(y+BB)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(*offset+m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(*offset+BB+*m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(**(*(*(&offset))+BB+*m)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*(x+*(y+*(**BB+BBB)+s7.i)+static_cast<int>(*(s7.s6[0].pp)))) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
#pragma omp target data map(to: s7.i)
{
#pragma omp target update from(s7.x)
}
+ const int &rci = i;
+ Base *b = new Base;
+ int *pd = const_cast<int*>(&d);
+#pragma omp target update from(const_cast<int&>(rci))
+#pragma omp target update from(const_cast<int*>(pd))
+#pragma omp target update from(static_cast<double>(i))
+#pragma omp target update from(dynamic_cast<Derived*>(b))
+#pragma omp target update from(reinterpret_cast<int*>(b))
+ {}
+
+#pragma omp target update from(static_cast<double>(i + 1)) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}} le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
+ {}
+
return tmain<int, 3>(argc)+tmain<to, 4>(argc); // expected-note {{in instantiation of function template specialization 'tmain<int, 3>' requested here}} expected-note {{in instantiation of function template specialization 'tmain<int, 4>' requested here}}
}
Index: clang/test/OpenMP/target_update_codegen.cpp
===================================================================
--- clang/test/OpenMP/target_update_codegen.cpp
+++ clang/test/OpenMP/target_update_codegen.cpp
@@ -291,5 +291,405 @@
#pragma omp target update from(arg) if(arg) device(4)
{++arg;}
}
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK5 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK5 --check-prefix CK5-64
+// RUN: %clang_cc1 -DCK5 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK5 --check-prefix CK5-64
+// RUN: %clang_cc1 -DCK5 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK5 --check-prefix CK5-32
+// RUN: %clang_cc1 -DCK5 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK5 --check-prefix CK5-32
+
+// RUN: %clang_cc1 -DCK5 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK5 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK5 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK5 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+
+#ifdef CK5
+
+// CK5: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
+// CK5: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
+
+// CK5-LABEL: lvalue
+void lvalue(int *B, int l, int e) {
+
+ // CK5-DAG: call void @__tgt_target_data_update(i64 -1, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK5-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK5-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+
+ // CK5-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
+ // CK5-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
+ // CK5-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to i32**
+ // CK5-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to i32**
+ // CK5-DAG: store i32* [[ZERO:%.+]], i32** [[BPC0]]
+ // CK5-DAG: store i32* [[ONE:%.+]], i32** [[PC0]]
+ // CK5-DAG: [[ZERO]] = load i32*, i32** [[B_ADDR:%.+]]
+ // CK5-DAG: [[ONE]] = load i32*, i32** [[B_ADDR]]
+ #pragma omp target update to(*B)
+ *B += e;
+ #pragma omp target update from(*B)
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK6 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK6 --check-prefix CK6-64
+// RUN: %clang_cc1 -DCK6 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK6 --check-prefix CK6-64
+// RUN: %clang_cc1 -DCK6 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK6 --check-prefix CK6-32
+// RUN: %clang_cc1 -DCK6 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK6 --check-prefix CK6-32
+
+// RUN: %clang_cc1 -DCK6 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK6 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK6 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK6 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+
+#ifdef CK6
+
+// CK6: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
+// CK6: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
+
+// CK6-LABEL: lvalue
+void lvalue(int *B, int l, int e) {
+
+ // CK6-DAG: call void @__tgt_target_data_update(i64 -1, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK6-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK6-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+
+ // CK6-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
+ // CK6-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
+ // CK6-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to i32**
+ // CK6-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to i32**
+ // CK6-DAG: store i32* [[ZERO:%.+]], i32** [[BPC0]]
+ // CK6-DAG: store i32* [[ADD_PTR:%.+]], i32** [[PC0]]
+ // CK6-DAG: [[ADD_PTR]] = getelementptr inbounds i32, i32* [[ONE:%.+]], i{{32|64}} [[IDX_EXT:%.+]]
+ // CK6-DAG: [[TWO:%.+]] = load i32, i32* [[L_ADDR:%.+]]
+ // CK6-DAG: store i32 [[L:%.+]], i32* [[L_ADDR]]
+ #pragma omp target update to(*(B+l))
+ *(B+l) += e;
+ #pragma omp target update from(*(B+l))
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK7 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK7 --check-prefix CK7-64
+// RUN: %clang_cc1 -DCK7 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK7 --check-prefix CK7-64
+// RUN: %clang_cc1 -DCK7 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK7 --check-prefix CK7-32
+// RUN: %clang_cc1 -DCK7 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK7 --check-prefix CK7-32
+
+// RUN: %clang_cc1 -DCK7 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK7 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK7 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK7 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+
+#ifdef CK7
+
+// CK7: [[SIZE00:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
+// CK7: [[MTYPE00:@.+]] = {{.+}}constant [1 x i64] [i64 33]
+
+// CK7-LABEL: lvalue
+void lvalue(int *B, int l, int e) {
+
+ // CK7-DAG: call void @__tgt_target_data_update(i64 -1, i32 1, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[1 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK7-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK7-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+
+ // CK7-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
+ // CK7-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
+ // CK7-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to i32**
+ // CK7-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to i32**
+ // CK7-DAG: store i32* [[ZERO:%.+]], i32** [[BPC0]]
+ // CK7-DAG: store i32* [[ARRAY_IDX:%.+]], i32** [[PC0]]
+ // CK7-DAG: [[ZERO]] = load i32*, i32** [[B_ADDR:%.+]]
+ // CK7-DAG: [[ARRAY_IDX]] = getelementptr inbounds i32, i32* [[ADD_PTR:%.+]], i{{32|64}} [[IDX_PROM:%.+]]
+ // CK7-64-DAG: [[ADD_PTR]] = getelementptr inbounds i32, i32* [[ONE:%.+]], i64 [[IDX_EXT:%.+]]
+ // CK7-32-DAG: [[ADD_PTR]] = getelementptr inbounds i32, i32* [[ONE:%.+]], i32 [[TWO:%.+]]
+ // CK7-32-DAG: [[ONE]] = load i32*, i32** [[B_ADDR]]
+ // CK7-32-DAG: [[TWO]] = load i32, i32* [[L_ADDR:%.+]]
+ // CK7-32-DAG: [[IDX_PROM]] = load i32, i32* [[L_ADDR]]
+ // CK7-64-DAG: [[IDX_EXT:%.+]] = sext i32 [[TWO:%.+]] to i64
+ // CK7-64-DAG: [[TWO:%.+]] = load i32, i32* [[L_ADDR:%.+]]
+ // CK7-64-DAG: [[IDX_PROM]] = sext i32 [[THREE:%.+]] to i64
+ // CK7-64-DAG: [[THREE]] = load i32, i32* [[L_ADDR]]
+ #pragma omp target update to((B+l)[l])
+ (B+l)[l] += e;
+ #pragma omp target update from((B+l)[l])
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK8 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK8 --check-prefix CK8-64
+// RUN: %clang_cc1 -DCK8 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK8 --check-prefix CK8-64
+// RUN: %clang_cc1 -DCK8 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK8 --check-prefix CK8-32
+// RUN: %clang_cc1 -DCK8 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK8 --check-prefix CK8-32
+
+// RUN: %clang_cc1 -DCK8 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK8 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK8 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK8 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+
+#ifdef CK8
+// CK8-64: [[SIZE00:@.+]] = {{.+}}constant [2 x i[[sz:64|32]]] [i{{64|32}} 8, i64 4]
+// CK8-32: [[SIZE00:@.+]] = {{.+}}constant [2 x i[[sz:64|32]]] [i{{64|32}} 4, i64 4]
+// CK8: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 33, i64 17]
+
+// CK8-LABEL: lvalue
+void lvalue(int **B, int l, int e) {
+
+ // CK8-DAG: call void @__tgt_target_data_update(i64 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}], [2 x i{{.+}}]* [[SIZE00]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}, [2 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK8-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK8-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+
+ // CK8-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
+ // CK8-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
+ // CK8-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to i32***
+ // CK8-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to i32**
+ // CK8-DAG: store i32** [[ARRAY_IDX_1:%.+]], i32*** [[BPC0]]
+ // CK8-DAG: store i32* [[ARRAY_IDX_4:%.+]], i32** [[PC0]]
+ // CK8-DAG: store i32** [[ARRAY_IDX_1]], i32*** [[NINE:%.+]]
+ // CK8-DAG: [[ARRAY_IDX_1]] = getelementptr inbounds i32*, i32** [[ADD_PTR:%.+]], i{{.+}} 1
+ // CK8-64-DAG: [[ADD_PTR]] = getelementptr inbounds i32*, i32** [[B_VAL:%.+]], i{{.+}} [[IDX_EXT:%.+]]
+ // CK8-64-DAG: [[IDX_EXT]] = sext i32 [[L_VAL:%.+]] to i64
+ // CK8-DAG: [[ARRAY_IDX_4]] = getelementptr inbounds i32, i32* [[FIVE:%.+]], i{{.+}} 2
+ // CK8-64-DAG: [[ARRAY_IDX_3:%.+]] = getelementptr inbounds i32*, i32** [[ADD_PTR_2:%.+]], i{{.+}} 1
+ // CK8-64-DAG: [[ADD_PTR_2]] = getelementptr inbounds i32*, i32** %3, i{{.+}} [[IDX_EXT_1:%.+]]
+ // CK8-64-DAG: [[IDX_EXT_1]] = sext i32 [[L_VAL:%.+]] to i{{.+}}
+ // CK8-32-DAG: [[ADD_PTR]] = getelementptr inbounds i32*, i32** [[B_VAL:%.+]], i{{.+}} [[L_VAL:%.+]]
+ // CK8-32-DAG: [[ARRAY_IDX_4:%.+]] = getelementptr inbounds i32*, i32** [[ADD_PTR_2:%.+]], i{{.+}} 1
+ // CK8-32-DAG: [[ADD_PTR_2]] = getelementptr inbounds i32*, i32** %3, i{{.+}} [[L_VAL:%.+]]
+ #pragma omp target update to((B+l)[1][2])
+ (B+l)[1][2] += e;
+ #pragma omp target update from((B+l)[1][2])
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK9 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK9 --check-prefix CK9-64
+// RUN: %clang_cc1 -DCK9 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK9 --check-prefix CK9-64
+// RUN: %clang_cc1 -DCK9 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK9 --check-prefix CK9-32
+// RUN: %clang_cc1 -DCK9 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK9 --check-prefix CK9-32
+
+// RUN: %clang_cc1 -DCK9 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK9 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK9 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK9 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+
+#ifdef CK9
+
+struct S {
+ double *p;
+};
+
+// CK9: [[SIZE00:@.+]] = {{.+}}constant [2 x i[[sz:64|32]]] [i{{64|32}} 32, i64 281474976710673]
+// CK9: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710674]
+
+// CK9-LABEL: lvalue
+void lvalue(struct S *s, int l, int e) {
+
+ // CK9-DAG: call void @__tgt_target_data_update(i64 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i{{.+}} [[GSIZE:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}, [2 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK9-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK9-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+ // CK9-DAG: [[GSIZE]] = getelementptr inbounds {{.+}}[[SIZE:%[^,]+]]
+ //
+ // CK9-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
+ // CK9-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
+ // CK9-DAG: [[SIZE0:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 1
+ // CK9-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to double***
+ // CK9-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to double**
+ // CK9-DAG: store double** [[P:%.+]], double*** [[BPC0]]
+ // CK9-DAG: store double* [[THREE:%.+]], double** [[PC0]]
+ // CK9-DAG: store i{{.+}} 8, i{{.+}}* [[SIZE0]]
+ // CK9-DAG: [[P]] = getelementptr inbounds [[STRUCT_S:%.+]], [[STRUCT_S]]* [[ONE:%.+]], i32 0, i32 0
+ // CK9-DAG: [[ONE]] = load [[STRUCT_S]]*, [[STRUCT_S]]** [[S_ADDR:%.+]]
+ // CK9-DAG: [[THREE]] = load double*, double** [[P_1:%.+]],
+ // CK9-DAG: [[P_1]] = getelementptr inbounds [[STRUCT_S]], [[STRUCT_S]]* [[TWO:%.+]], i32 0, i32 0
+ // CK9-DAG: [[TWO]] = load [[STRUCT_S]]*, [[STRUCT_S]]** [[S_ADDR:%.+]]
+ #pragma omp target update to(*(s->p))
+ *(s->p) += e;
+ #pragma omp target update from(*(s->p))
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK10 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK10 --check-prefix CK10-64
+// RUN: %clang_cc1 -DCK10 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK10 --check-prefix CK10-64
+// RUN: %clang_cc1 -DCK10 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK10 --check-prefix CK10-32
+// RUN: %clang_cc1 -DCK10 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK10 --check-prefix CK10-32
+
+// RUN: %clang_cc1 -DCK10 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK10 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK10 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK10 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+#ifdef CK10
+
+struct S {
+ double *p;
+};
+
+// CK10: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710674]
+
+// CK10-LABEL: lvalue
+void lvalue(struct S *s, int l, int e) {
+
+ // CK10-DAG: call void @__tgt_target_data_update(i64 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i{{.+}} [[GSIZE:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}, [2 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK10-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK10-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+ // CK10-DAG: [[GSIZE]] = getelementptr inbounds {{.+}}[[SIZE:%[^,]+]]
+ //
+ // CK10-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
+ // CK10-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
+ // CK10-DAG: [[SIZE0:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 1
+ // CK10-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to double***
+ // CK10-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to double**
+ // CK10-DAG: store double** [[P_VAL:%.+]], double*** [[BPC0]]
+ // CK10-DAG: store double* [[ADD_PTR:%.+]], double** [[PC0]]
+ // CK10-DAG: store i{{.+}} 8, i{{.+}}* [[SIZE0]]
+ // CK10-64-DAG: [[ADD_PTR]] = getelementptr inbounds double, double* [[THREE:%.+]], i{{.+}} [[IDX_EXT:%.+]]
+ // CK10-32-DAG: [[ADD_PTR]] = getelementptr inbounds double, double* [[THREE:%.+]], i{{.+}} [[L_VAL:%.+]]
+ // CK10-64-DAG: [[IDX_EXT]] = sext i32 [[L_VAL:%.+]] to i64
+ // CK10-DAG: [[THREE]] = load double*, double** [[P_VAL_1:%.+]]
+ // CK10-DAG: getelementptr inbounds {{.+}}, {{.+}}* [[SS_1:%.+]], i32 0, i32 0
+ // CK10-DAG: [[SS_1]] = load {{.+}}*, {{.+}}** [[S_ADDR:%.+]]
+ #pragma omp target update to(*(s->p+l))
+ *(s->p+l) += e;
+ #pragma omp target update from(*(s->p+l))
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK11 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK11 --check-prefix CK11-64
+// RUN: %clang_cc1 -DCK11 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK11 --check-prefix CK11-64
+// RUN: %clang_cc1 -DCK11 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK11 --check-prefix CK11-32
+// RUN: %clang_cc1 -DCK11 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK11 --check-prefix CK11-32
+
+// RUN: %clang_cc1 -DCK11 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK11 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK11 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK11 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+#ifdef CK11
+
+struct S {
+ double *p;
+};
+// CK11: [[MTYPE00:@.+]] = {{.+}}constant [2 x i64] [i64 32, i64 281474976710674]
+
+// CK11-LABEL: lvalue
+void lvalue(struct S *s, int l, int e) {
+
+ // CK11-DAG: call void @__tgt_target_data_update(i64 -1, i32 2, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i{{.+}} [[GSIZE:%.+]], {{.+}}getelementptr {{.+}}[2 x i{{.+}}, [2 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK11-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK11-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+ // CK11-DAG: [[GSIZE]] = getelementptr inbounds {{.+}}[[SIZE:%[^,]+]]
+ //
+ // CK11-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
+ // CK11-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
+ // CK11-DAG: [[SIZE0:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 1
+ // CK11-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to double***
+ // CK11-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to double**
+ // CK11-DAG: store double** [[P:%.+]], double*** [[BPC0]]
+ // CK11-DAG: store double* [[ARRAY_IDX:%.+]], double** [[PC0]]
+ // CK11-DAG: store i{{.+}} 8, i{{.+}} [[SIZE0]]
+ // CK11-DAG: [[P]] = getelementptr inbounds [[STRUCT_S:%.+]], [[STRUCT_S]]* [[SS_1:%.+]], i32 0, i32 0
+ // CK11-DAG: [[ARRAY_IDX]] = getelementptr inbounds double, double* [[ADD_PTR:%.+]], i{{.+}} 3
+ // CK11-64-DAG: [[ADD_PTR]] = getelementptr inbounds double, double* [[THREE:%.+]], i{{.+}} [[IDX_EXT:%.+]]
+ // CK11-32-DAG: [[ADD_PTR]] = getelementptr inbounds double, double* [[THREE:%.+]], i{{.+}} [[L_VAL:%.+]]
+ // CK11-64-DAG: [[IDX_EXT]] = sext i32 [[L_VAL:%.+]] to i64
+ #pragma omp target update to((s->p+l)[3])
+ (s->p+l)[3] += e;
+ #pragma omp target update from((s->p+l)[3])
+}
+
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK12 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK12 --check-prefix CK12-64
+// RUN: %clang_cc1 -DCK12 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK12 --check-prefix CK12-64
+// RUN: %clang_cc1 -DCK12 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck %s --check-prefix CK12 --check-prefix CK12-32
+// RUN: %clang_cc1 -DCK12 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CK12 --check-prefix CK12-32
+
+// RUN: %clang_cc1 -DCK12 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK12 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK12 -verify -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// RUN: %clang_cc1 -DCK12 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s
+// SIMD-ONLY0-NOT: {{__kmpc|__tgt}}
+#ifdef CK12
+
+struct S {
+ double *p;
+ struct S *sp;
+};
+// CK12: [[MTYPE00:@.+]] = {{.+}}constant [3 x i64] [i64 32, i64 281474976710672, i64 17]
+
+// CK12-LABEL: lvalue
+void lvalue(struct S *s, int l, int e) {
+
+ // CK12-DAG: call void @__tgt_target_data_update(i64 -1, i32 3, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], i{{.+}} [[GSIZE:%.+]], {{.+}}getelementptr {{.+}}[3 x i{{.+}}, [3 x i{{.+}}]* [[MTYPE00]]{{.+}})
+ // CK12-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
+ // CK12-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
+ // CK12-DAG: [[GSIZE]] = getelementptr inbounds {{.+}}[[SIZE:%[^,]+]]
+ //
+ // CK12-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 2
+ // CK12-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 2
+ // CK12-DAG: [[SIZE2:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 2
+ // CK12-DAG: [[BPC2:%.+]] = bitcast i8** [[BP2]] to double***
+ // CK12-DAG: [[PC2:%.+]] = bitcast i8** [[P2]] to double**
+ // CK12-DAG: store double** [[P_VAL:%.+]], double*** [[BPC2]]
+ // CK12-DAG: store double* [[SIX:%.+]], double** [[PC2]]
+ // CK12-DAG: store i{{.+}} 8, i{{.+}}* [[SIZE2]]
+ // CK12-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
+ // CK12-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
+ // CK12-DAG: [[SIZE1:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 1
+ // CK12-DAG: [[BPC1:%.+]] = bitcast i8** [[BP1]] to [[STRUCT_S:%.+]]***
+ // CK12-DAG: [[PC1:%.+]] = bitcast i8** [[P1]] to double***
+ // CK12-DAG: store [[STRUCT_S]]** [[SP:%.+]], [[STRUCT_S]]*** [[BPC1]]
+ // CK12-DAG: store double** [[P_VAL:%.+]], double*** [[PC1]]
+ // CK12-DAG: store i{{.+}} {{4|8}}, i{{.+}}* [[SIZE1]]
+ // CK12-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
+ // CK12-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
+ // CK12-DAG: [[SIZE0:%.+]] = getelementptr inbounds {{.+}}[[SIZE]], i{{.+}} 0, i{{.+}} 0
+ // CK12-DAG: [[BPC0:%.+]] = bitcast i8** [[BP0]] to [[STRUCT_S:%.+]]**
+ // CK12-DAG: [[PC0:%.+]] = bitcast i8** [[P0]] to [[STRUCT_S]]***
+ // CK12-DAG: store [[STRUCT_S]]** [[SP:%.+]], [[STRUCT_S]]*** [[S_VAL:%.+]]
+ // CK12-DAG: store i{{.+}} {{.+}}, i{{.+}}* [[SIZE0]]
+ // CK12-DAG: [[SP]] = getelementptr inbounds [[STRUCT_S]], [[STRUCT_S]]* [[ONE:%.+]], i32 0, i32 1
+ #pragma omp target update to(*(s->sp->p))
+ *(s->sp->p) = e;
+ #pragma omp target update from(*(s->sp->p))
+}
+
#endif
#endif
Index: clang/test/OpenMP/target_teams_map_messages.cpp
===================================================================
--- clang/test/OpenMP/target_teams_map_messages.cpp
+++ clang/test/OpenMP/target_teams_map_messages.cpp
@@ -1,10 +1,14 @@
// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=40 -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-version=45 -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -verify=expected -fopenmp-version=50 -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 200 -x c %s -Wno-openmp -Wuninitialized
// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
-// RUN: %clang_cc1 -DCCODE -verify=expected,le45 -fopenmp -ferror-limit 200 -x c %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-simd -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -DCCODE -verify -fopenmp-simd -ferror-limit 200 -x c %s -Wno-openmp-mapping -Wuninitialized
#ifdef CCODE
void foo(int arg) {
const int n = 0;
@@ -44,9 +48,9 @@
{}
#pragma omp target teams map(arg[2:2],a,d) // expected-error {{subscripted value is not an array or pointer}}
{}
- #pragma omp target teams map(arg,a*2) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target teams map(arg,a*2) // le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
- #pragma omp target teams map(arg,(c+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target teams map(arg,(c+1)[2]) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
#pragma omp target teams map(arg,a[:2],d) // expected-error {{subscripted value is not an array or pointer}}
{}
@@ -252,7 +256,7 @@
{}
#pragma omp target teams map(r.S.Arr[:12])
{}
- #pragma omp target teams map(r.S.foo()[:12]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target teams map(r.S.foo()[:12]) // le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
#pragma omp target teams map(r.C, r.D)
{}
@@ -286,7 +290,7 @@
{}
#pragma omp target teams map(r.S.Ptr[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
{}
- #pragma omp target teams map((p+1)->A) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target teams map((p+1)->A) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
#pragma omp target teams map(u.B) // expected-error {{mapping of union members is not allowed}}
{}
@@ -416,7 +420,7 @@
foo();
#pragma omp target teams map(T) // expected-error {{'T' does not refer to a value}}
foo();
-#pragma omp target teams map(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams map(I) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
foo();
#pragma omp target teams map(S2::S2s)
foo();
@@ -434,7 +438,8 @@
foo();
#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target data map(tofrom: argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: argc > 0 ? x : y) // le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+
#pragma omp target data map(argc)
#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
@@ -508,7 +513,7 @@
foo();
#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // expected-error {{xpected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}} le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
#pragma omp target data map(argc)
#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
Index: clang/test/OpenMP/target_map_messages.cpp
===================================================================
--- clang/test/OpenMP/target_map_messages.cpp
+++ clang/test/OpenMP/target_map_messages.cpp
@@ -1,7 +1,13 @@
-// RUN: %clang_cc1 -verify -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
// RUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 200 -x c %s -Wno-openmp -Wuninitialized
-// RUN: %clang_cc1 -verify -fopenmp-simd -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le45 -fopenmp-simd -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-simd -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
// RUN: %clang_cc1 -DCCODE -verify -fopenmp-simd -ferror-limit 200 -x c %s -Wno-openmp-mapping -Wuninitialized
#ifdef CCODE
void foo(int arg) {
@@ -57,9 +63,9 @@
{}
#pragma omp target map(arg[2:2],a,d) // expected-error {{subscripted value is not an array or pointer}}
{}
- #pragma omp target map(arg,a*2) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target map(arg,a*2) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
{}
- #pragma omp target map(arg,(c+1)[2]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target map(arg,(c+1)[2]) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
#pragma omp target map(arg,a[:2],d) // expected-error {{subscripted value is not an array or pointer}}
{}
@@ -305,7 +311,7 @@
{}
#pragma omp target map(r.S.Arr[:12])
{}
- #pragma omp target map(r.S.foo()[:12]) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target map(r.S.foo()[:12]) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
{}
#pragma omp target map(r.C, r.D)
{}
@@ -339,7 +345,7 @@
{}
#pragma omp target map(r.S.Ptr[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
{}
- #pragma omp target map((p+1)->A) // expected-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+ #pragma omp target map((p+1)->A) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}}
{}
#pragma omp target map(u.B) // expected-error {{mapping of union members is not allowed}}
{}
@@ -479,7 +485,7 @@
foo();
#pragma omp target map(T) // expected-error {{'T' does not refer to a value}}
foo();
-#pragma omp target map(I) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target map(I) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}} le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
foo();
#pragma omp target map(S2::S2s)
foo();
@@ -496,7 +502,7 @@
#pragma omp target map(to, x)
foo();
#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target data map(tofrom: argc > 0 ? x : y) // expected-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: argc > 0 ? x : y) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}} le50-error 2 {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
#pragma omp target data map(argc)
#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
@@ -592,7 +598,7 @@
#pragma omp target map(to, x)
foo();
#pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // expected-error {{xpected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: argc > 0 ? argv[1] : argv[2]) // le45-error {{expected expression containing only member accesses and/or array sections based on named variables}} le50-error {{expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'}}
#pragma omp target data map(argc)
#pragma omp target data map(S1) // expected-error {{'S1' does not refer to a value}}
#pragma omp target data map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -15197,256 +15197,411 @@
return ConstLength.getSExtValue() != 1;
}
-// Return the expression of the base of the mappable expression or null if it
-// cannot be determined and do all the necessary checks to see if the expression
-// is valid as a standalone mappable expression. In the process, record all the
-// components of the expression.
-static const Expr *checkMapClauseExpressionBase(
- Sema &SemaRef, Expr *E,
- OMPClauseMappableExprCommon::MappableExprComponentList &CurComponents,
- OpenMPClauseKind CKind, bool NoDiagnose) {
- SourceLocation ELoc = E->getExprLoc();
- SourceRange ERange = E->getSourceRange();
-
- // The base of elements of list in a map clause have to be either:
- // - a reference to variable or field.
- // - a member expression.
- // - an array expression.
- //
- // E.g. if we have the expression 'r.S.Arr[:12]', we want to retrieve the
- // reference to 'r'.
- //
- // If we have:
- //
- // struct SS {
- // Bla S;
- // foo() {
- // #pragma omp target map (S.Arr[:12]);
- // }
- // }
- //
- // We want to retrieve the member expression 'this->S';
-
- const Expr *RelevantExpr = nullptr;
-
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.2]
- // If a list item is an array section, it must specify contiguous storage.
- //
- // For this restriction it is sufficient that we make sure only references
- // to variables or fields and array expressions, and that no array sections
- // exist except in the rightmost expression (unless they cover the whole
- // dimension of the array). E.g. these would be invalid:
- //
- // r.ArrS[3:5].Arr[6:7]
- //
- // r.ArrS[3:5].x
- //
- // but these would be valid:
- // r.ArrS[3].Arr[6:7]
- //
- // r.ArrS[3].x
+namespace {
+// The base of elements of list in a map clause have to be either:
+// - a reference to variable or field.
+// - a member expression.
+// - an array expression.
+//
+// E.g. if we have the expression 'r.S.Arr[:12]', we want to retrieve the
+// reference to 'r'.
+//
+// If we have:
+//
+// struct SS {
+// Bla S;
+// foo() {
+// #pragma omp target map (S.Arr[:12]);
+// }
+// }
+//
+// We want to retrieve the member expression 'this->S';
- bool AllowUnitySizeArraySection = true;
- bool AllowWholeSizeArraySection = true;
+// OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.2]
+// If a list item is an array section, it must specify contiguous storage.
+//
+// For this restriction it is sufficient that we make sure only references
+// to variables or fields and array expressions, and that no array sections
+// exist except in the rightmost expression (unless they cover the whole
+// dimension of the array). E.g. these would be invalid:
+//
+// r.ArrS[3:5].Arr[6:7]
+//
+// r.ArrS[3:5].x
+//
+// but these would be valid:
+// r.ArrS[3].Arr[6:7]
+//
+// r.ArrS[3].x
+class MapBaseChecker final : public StmtVisitor<MapBaseChecker, bool> {
+ Sema &SemaRef;
+ OpenMPClauseKind CKind;
+ OMPClauseMappableExprCommon::MappableExprComponentList &Components;
+ bool NoDiagnose;
+ const Expr *RelevantExpr{nullptr};
+ size_t Depth{0};
+ SmallVector<int, 4> CntStack;
+ bool AllowUnitySizeArraySection{true};
+ bool AllowWholeSizeArraySection{true};
+ SourceLocation ELoc;
+ SourceRange ERange;
+
+ void emitErrorMsg() {
+ if (!NoDiagnose) {
+ // If nothing else worked, this is not a valid map clause expression.
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ SemaRef.Diag(
+ ELoc, diag::err_omp_expected_named_var_member_or_array_expression)
+ << ERange;
+ } else {
+ SemaRef.Diag(ELoc, diag::err_omp_non_lvalue_in_map_or_motion_clauses)
+ << ERange;
+ }
+ }
+ }
- while (!RelevantExpr) {
- E = E->IgnoreParenImpCasts();
+public:
+ bool VisitDeclRefExpr(DeclRefExpr *DRE) {
+ if (!isa<VarDecl>(DRE->getDecl()))
+ return false;
- if (auto *CurE = dyn_cast<DeclRefExpr>(E)) {
- if (!isa<VarDecl>(CurE->getDecl()))
- return nullptr;
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ RelevantExpr = DRE;
+ // Record the component.
+ Components.emplace_back(DRE, DRE->getDecl());
+ } else {
+ const clang::Type *PT = DRE->getType().getTypePtr();
+ const std::string TypeStr = PT->getCanonicalTypeInternal().getAsString();
+ size_t TypeDecorCnt = std::count_if(TypeStr.begin(), TypeStr.end(),
+ [](char c) { return c == '*' || c == '['; });
+ // Normally we want to add Components if Depth == TypeDecorCnt, for example:
+ //
+ // int **B, *l;
+ // map(**(B + *l))
+ // Depth for B: 2
+ // TypeDecorCnt for B: 2
+ // Depth for l: 3
+ // TypeDecorCnt for l: 1
+ //
+ // However, combining with address_of operator(&), we might have Depth
+ // less than TypeDecorCnt as LValue
+ //
+ // int B;
+ // map(*(&B))
+ // Depth for B: 1
+ // TypeDecorCnt for B: 0
+ if (Depth <= TypeDecorCnt) {
+ RelevantExpr = cast<Expr>(DRE);
+ // record the component.
+ Components.emplace_back(DRE, DRE->getDecl());
+ }
+ }
- RelevantExpr = CurE;
+ return true;
+ }
+ bool VisitMemberExpr(MemberExpr *ME) {
+ Expr *E = cast<Expr>(ME);
+ Expr *BaseE = ME->getBase()->IgnoreParenImpCasts();
+ CntStack.emplace_back(Depth);
+ Depth = 0;
- // If we got a reference to a declaration, we should not expect any array
- // section before that.
- AllowUnitySizeArraySection = false;
- AllowWholeSizeArraySection = false;
+ if (isa<CXXThisExpr>(BaseE))
+ // We found a base expression: this->Val.
+ RelevantExpr = ME;
+ else
+ E = BaseE;
- // Record the component.
- CurComponents.emplace_back(CurE, CurE->getDecl());
- } else if (auto *CurE = dyn_cast<MemberExpr>(E)) {
- Expr *BaseE = CurE->getBase()->IgnoreParenImpCasts();
+ if (!isa<FieldDecl>(ME->getMemberDecl())) {
+ if (!NoDiagnose) {
+ SemaRef.Diag(ELoc, diag::err_omp_expected_access_to_data_field)
+ << ME->getSourceRange();
+ return false;
+ }
+ if (RelevantExpr)
+ return false;
+ return Visit(E->IgnoreParenImpCasts());;
+ }
- if (isa<CXXThisExpr>(BaseE))
- // We found a base expression: this->Val.
- RelevantExpr = CurE;
- else
- E = BaseE;
+ auto *FD = cast<FieldDecl>(ME->getMemberDecl());
- if (!isa<FieldDecl>(CurE->getMemberDecl())) {
- if (!NoDiagnose) {
- SemaRef.Diag(ELoc, diag::err_omp_expected_access_to_data_field)
- << CurE->getSourceRange();
- return nullptr;
- }
- if (RelevantExpr)
- return nullptr;
- continue;
+ // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.3]
+ // A bit-field cannot appear in a map clause.
+ //
+ if (FD->isBitField()) {
+ if (!NoDiagnose) {
+ SemaRef.Diag(ELoc, diag::err_omp_bit_fields_forbidden_in_clause)
+ << ME->getSourceRange() << getOpenMPClauseName(CKind);
+ return false;
}
+ if (RelevantExpr)
+ return false;
+ return Visit(E->IgnoreParenImpCasts());;
+ }
- auto *FD = cast<FieldDecl>(CurE->getMemberDecl());
+ // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C++, p.1]
+ // If the type of a list item is a reference to a type T then the type
+ // will be considered to be T for all purposes of this clause.
+ QualType CurType = BaseE->getType().getNonReferenceType();
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.3]
- // A bit-field cannot appear in a map clause.
- //
- if (FD->isBitField()) {
- if (!NoDiagnose) {
- SemaRef.Diag(ELoc, diag::err_omp_bit_fields_forbidden_in_clause)
- << CurE->getSourceRange() << getOpenMPClauseName(CKind);
- return nullptr;
- }
- if (RelevantExpr)
- return nullptr;
- continue;
+ // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.2]
+ // A list item cannot be a variable that is a member of a structure with
+ // a union type.
+ //
+ if (CurType->isUnionType()) {
+ if (!NoDiagnose) {
+ SemaRef.Diag(ELoc, diag::err_omp_union_type_not_allowed)
+ << ME->getSourceRange();
+ return false;
}
+ return Visit(E->IgnoreParenImpCasts());;
+ }
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C++, p.1]
- // If the type of a list item is a reference to a type T then the type
- // will be considered to be T for all purposes of this clause.
- QualType CurType = BaseE->getType().getNonReferenceType();
+ // If we got a member expression, we should not expect any array section
+ // before that:
+ //
+ // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.7]
+ // If a list item is an element of a structure, only the rightmost symbol
+ // of the variable reference can be an array section.
+ //
+ AllowUnitySizeArraySection = false;
+ AllowWholeSizeArraySection = false;
+
+ // Record the component.
+ Components.emplace_back(ME, FD);
+ if (RelevantExpr || Visit(E->IgnoreParenImpCasts())) {
+ Depth = CntStack.back();
+ CntStack.pop_back();
+ return true;
+ } else {
+ return false;
+ }
+ }
+ bool VisitCXXThisExpr(CXXThisExpr *CTE) {
+ if (!RelevantExpr) {
+ Components.emplace_back(CTE, nullptr);
+ this->RelevantExpr = cast<Expr>(CTE);
+ }
+ return true;
+ }
+ bool VisitUnaryOperator(UnaryOperator *UO) {
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ emitErrorMsg();
+ return false;
+ }
+ if (UO->getOpcode() == UO_Deref || UO->getOpcode() == UO_AddrOf) {
+ if (!this->RelevantExpr && UO->getOpcode() == UO_Deref) {
+ Depth++;
+ // Record the component.
+ Components.emplace_back(UO, nullptr);
+ } else {
+ Depth--;
+ }
+ Expr *SubE = UO->getSubExpr()->IgnoreParenImpCasts();
+ if (SubE && Visit(SubE->IgnoreParenImpCasts())) {
+ if (UO->getOpcode() == UO_Deref)
+ Depth--;
+ else
+ Depth++;
+ return true;
+ }
+ }
+ return false;
+ }
+ bool VisitArraySubscriptExpr(ArraySubscriptExpr *AE) {
+ Depth++;
+ Expr *E = AE->getBase()->IgnoreParenImpCasts();
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.2]
- // A list item cannot be a variable that is a member of a structure with
- // a union type.
- //
- if (CurType->isUnionType()) {
- if (!NoDiagnose) {
- SemaRef.Diag(ELoc, diag::err_omp_union_type_not_allowed)
- << CurE->getSourceRange();
- return nullptr;
- }
- continue;
+ if (!E->getType()->isAnyPointerType() && !E->getType()->isArrayType()) {
+ if (!NoDiagnose) {
+ SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
+ << 0 << AE->getSourceRange();
+ return false;
}
+ return Visit(E->IgnoreParenImpCasts());
+ }
- // If we got a member expression, we should not expect any array section
- // before that:
- //
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.7]
- // If a list item is an element of a structure, only the rightmost symbol
- // of the variable reference can be an array section.
- //
- AllowUnitySizeArraySection = false;
+ // If we got an array subscript that express the whole dimension we
+ // can have any array expressions before. If it only expressing part of
+ // the dimension, we can only have unitary-size array expressions.
+ if (checkArrayExpressionDoesNotReferToWholeSize(SemaRef, AE,
+ E->getType()))
AllowWholeSizeArraySection = false;
- // Record the component.
- CurComponents.emplace_back(CurE, FD);
- } else if (auto *CurE = dyn_cast<ArraySubscriptExpr>(E)) {
- E = CurE->getBase()->IgnoreParenImpCasts();
-
- if (!E->getType()->isAnyPointerType() && !E->getType()->isArrayType()) {
- if (!NoDiagnose) {
- SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
- << 0 << CurE->getSourceRange();
- return nullptr;
+ if (const auto *TE = dyn_cast<CXXThisExpr>(E)) {
+ Expr::EvalResult Result;
+ if (AE->getIdx()->EvaluateAsInt(Result, SemaRef.getASTContext())) {
+ if (!Result.Val.getInt().isNullValue()) {
+ SemaRef.Diag(AE->getIdx()->getExprLoc(),
+ diag::err_omp_invalid_map_this_expr);
+ SemaRef.Diag(AE->getIdx()->getExprLoc(),
+ diag::note_omp_invalid_subscript_on_this_ptr_map);
}
- continue;
}
+ RelevantExpr = TE;
+ }
- // If we got an array subscript that express the whole dimension we
- // can have any array expressions before. If it only expressing part of
- // the dimension, we can only have unitary-size array expressions.
- if (checkArrayExpressionDoesNotReferToWholeSize(SemaRef, CurE,
- E->getType()))
- AllowWholeSizeArraySection = false;
-
- if (const auto *TE = dyn_cast<CXXThisExpr>(E)) {
- Expr::EvalResult Result;
- if (CurE->getIdx()->EvaluateAsInt(Result, SemaRef.getASTContext())) {
- if (!Result.Val.getInt().isNullValue()) {
- SemaRef.Diag(CurE->getIdx()->getExprLoc(),
- diag::err_omp_invalid_map_this_expr);
- SemaRef.Diag(CurE->getIdx()->getExprLoc(),
- diag::note_omp_invalid_subscript_on_this_ptr_map);
- }
- }
- RelevantExpr = TE;
- }
+ // Record the component - we don't have any declaration associated.
+ Components.emplace_back(AE, nullptr);
- // Record the component - we don't have any declaration associated.
- CurComponents.emplace_back(CurE, nullptr);
- } else if (auto *CurE = dyn_cast<OMPArraySectionExpr>(E)) {
- assert(!NoDiagnose && "Array sections cannot be implicitly mapped.");
- E = CurE->getBase()->IgnoreParenImpCasts();
+ if (E && Visit(E->IgnoreParenImpCasts())) {
+ Depth--;
+ return true;
+ }
+ return false;
+ }
+ bool VisitOMPArraySectionExpr(OMPArraySectionExpr *OASE) {
+ assert(!NoDiagnose && "Array sections cannot be implicitly mapped.");
+ Depth++;
+ Expr *E = OASE->getBase()->IgnoreParenImpCasts();
- QualType CurType =
- OMPArraySectionExpr::getBaseOriginalType(E).getCanonicalType();
+ QualType CurType =
+ OMPArraySectionExpr::getBaseOriginalType(E).getCanonicalType();
- // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C++, p.1]
- // If the type of a list item is a reference to a type T then the type
- // will be considered to be T for all purposes of this clause.
- if (CurType->isReferenceType())
- CurType = CurType->getPointeeType();
+ // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C++, p.1]
+ // If the type of a list item is a reference to a type T then the type
+ // will be considered to be T for all purposes of this clause.
+ if (CurType->isReferenceType())
+ CurType = CurType->getPointeeType();
- bool IsPointer = CurType->isAnyPointerType();
+ bool IsPointer = CurType->isAnyPointerType();
- if (!IsPointer && !CurType->isArrayType()) {
- SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
- << 0 << CurE->getSourceRange();
- return nullptr;
- }
+ if (!IsPointer && !CurType->isArrayType()) {
+ SemaRef.Diag(ELoc, diag::err_omp_expected_base_var_name)
+ << 0 << OASE->getSourceRange();
+ return false;
+ }
- bool NotWhole =
- checkArrayExpressionDoesNotReferToWholeSize(SemaRef, CurE, CurType);
- bool NotUnity =
- checkArrayExpressionDoesNotReferToUnitySize(SemaRef, CurE, CurType);
+ bool NotWhole =
+ checkArrayExpressionDoesNotReferToWholeSize(SemaRef, OASE, CurType);
+ bool NotUnity =
+ checkArrayExpressionDoesNotReferToUnitySize(SemaRef, OASE, CurType);
- if (AllowWholeSizeArraySection) {
- // Any array section is currently allowed. Allowing a whole size array
- // section implies allowing a unity array section as well.
- //
- // If this array section refers to the whole dimension we can still
- // accept other array sections before this one, except if the base is a
- // pointer. Otherwise, only unitary sections are accepted.
- if (NotWhole || IsPointer)
- AllowWholeSizeArraySection = false;
- } else if (AllowUnitySizeArraySection && NotUnity) {
- // A unity or whole array section is not allowed and that is not
- // compatible with the properties of the current array section.
- SemaRef.Diag(
- ELoc, diag::err_array_section_does_not_specify_contiguous_storage)
- << CurE->getSourceRange();
- return nullptr;
- }
+ if (AllowWholeSizeArraySection) {
+ // Any array section is currently allowed. Allowing a whole size array
+ // section implies allowing a unity array section as well.
+ //
+ // If this array section refers to the whole dimension we can still
+ // accept other array sections before this one, except if the base is a
+ // pointer. Otherwise, only unitary sections are accepted.
+ if (NotWhole || IsPointer)
+ AllowWholeSizeArraySection = false;
+ } else if (AllowUnitySizeArraySection && NotUnity) {
+ // A unity or whole array section is not allowed and that is not
+ // compatible with the properties of the current array section.
+ SemaRef.Diag(
+ ELoc, diag::err_array_section_does_not_specify_contiguous_storage)
+ << OASE->getSourceRange();
+ return false;
+ }
- if (const auto *TE = dyn_cast<CXXThisExpr>(E)) {
- Expr::EvalResult ResultR;
- Expr::EvalResult ResultL;
- if (CurE->getLength()->EvaluateAsInt(ResultR,
- SemaRef.getASTContext())) {
- if (!ResultR.Val.getInt().isOneValue()) {
- SemaRef.Diag(CurE->getLength()->getExprLoc(),
- diag::err_omp_invalid_map_this_expr);
- SemaRef.Diag(CurE->getLength()->getExprLoc(),
- diag::note_omp_invalid_length_on_this_ptr_mapping);
- }
+ if (const auto *TE = dyn_cast<CXXThisExpr>(E)) {
+ Expr::EvalResult ResultR;
+ Expr::EvalResult ResultL;
+ if (OASE->getLength()->EvaluateAsInt(ResultR,
+ SemaRef.getASTContext())) {
+ if (!ResultR.Val.getInt().isOneValue()) {
+ SemaRef.Diag(OASE->getLength()->getExprLoc(),
+ diag::err_omp_invalid_map_this_expr);
+ SemaRef.Diag(OASE->getLength()->getExprLoc(),
+ diag::note_omp_invalid_length_on_this_ptr_mapping);
}
- if (CurE->getLowerBound() && CurE->getLowerBound()->EvaluateAsInt(
- ResultL, SemaRef.getASTContext())) {
- if (!ResultL.Val.getInt().isNullValue()) {
- SemaRef.Diag(CurE->getLowerBound()->getExprLoc(),
- diag::err_omp_invalid_map_this_expr);
- SemaRef.Diag(CurE->getLowerBound()->getExprLoc(),
- diag::note_omp_invalid_lower_bound_on_this_ptr_mapping);
- }
+ }
+ if (OASE->getLowerBound() && OASE->getLowerBound()->EvaluateAsInt(
+ ResultL, SemaRef.getASTContext())) {
+ if (!ResultL.Val.getInt().isNullValue()) {
+ SemaRef.Diag(OASE->getLowerBound()->getExprLoc(),
+ diag::err_omp_invalid_map_this_expr);
+ SemaRef.Diag(OASE->getLowerBound()->getExprLoc(),
+ diag::note_omp_invalid_lower_bound_on_this_ptr_mapping);
}
- RelevantExpr = TE;
}
+ RelevantExpr = TE;
+ }
- // Record the component - we don't have any declaration associated.
- CurComponents.emplace_back(CurE, nullptr);
- } else {
- if (!NoDiagnose) {
- // If nothing else worked, this is not a valid map clause expression.
- SemaRef.Diag(
- ELoc, diag::err_omp_expected_named_var_member_or_array_expression)
- << ERange;
- }
- return nullptr;
+ // Record the component - we don't have any declaration associated.
+ Components.emplace_back(OASE, nullptr);
+
+ if (RelevantExpr || Visit(E->IgnoreParenImpCasts())) {
+ Depth--;
+ return true;
+ }
+ return false;
+ }
+ bool VisitBinaryOperator(BinaryOperator *BO) {
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ emitErrorMsg();
+ return false;
+ }
+ if (BO->getOpcode() == BO_Assign) {
+ this->RelevantExpr = cast<Expr>(BO);
+ // Record the component.
+ Components.emplace_back(BO, nullptr);
+ return true;
}
+ Expr *LE = BO->getLHS()->IgnoreParenImpCasts();
+ Expr *RE = BO->getRHS()->IgnoreParenImpCasts();
+ if (!LE && !RE)
+ return false;
+ if (LE && RE) {
+ Visit(LE);
+ Visit(RE);
+ return true;
+ }
+ return false;
+ }
+ bool VisitConditionalOperator(ConditionalOperator *CO) {
+ emitErrorMsg();
+ return false;
}
+ bool VisitCallExpr(CallExpr *CE) {
+ emitErrorMsg();
+ return false;
+ }
+ bool VisitCastExpr(CastExpr *CE) {
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ emitErrorMsg();
+ return false;
+ }
+ if (!CE->isLValue())
+ return false;
+ return Visit(CE->getSubExpr());
+ }
+ bool VisitCXXMemberCallExpr(CXXMemberCallExpr *CCE) {
+ emitErrorMsg();
+ return false;
+ }
+ const Expr *getFoundBase() const {
+ return RelevantExpr;
+ }
+ OMPClauseMappableExprCommon::MappableExprComponentList getComponents() {
+ return Components;
+ }
+ explicit MapBaseChecker(
+ Sema &SemaRef, OpenMPClauseKind CKind,
+ OMPClauseMappableExprCommon::MappableExprComponentList &Components,
+ bool NoDiagnose, SourceLocation &ELoc, SourceRange &ERange)
+ : SemaRef(SemaRef), CKind(CKind), Components(Components),
+ NoDiagnose(NoDiagnose), ELoc(ELoc), ERange(ERange) {}
+};
+} // namespace
- return RelevantExpr;
+// Return the expression of the base of the mappable expression or null if it
+// cannot be determined and do all the necessary checks to see if the expression
+// is valid as a standalone mappable expression. In the process, record all the
+// components of the expression.
+static const Expr *checkMapClauseExpressionBase(
+ Sema &SemaRef, Expr *E,
+ OMPClauseMappableExprCommon::MappableExprComponentList &CurComponents,
+ OpenMPClauseKind CKind, bool NoDiagnose) {
+ SourceLocation ELoc = E->getExprLoc();
+ SourceRange ERange = E->getSourceRange();
+ MapBaseChecker Checker(SemaRef, CKind, CurComponents, NoDiagnose, ELoc,
+ ERange);
+ if (Checker.Visit(E->IgnoreParenImpCasts())) {
+ return Checker.getFoundBase();
+ } else {
+ return nullptr;
+ }
}
// Return true if expression E associated with value VD has conflicts with other
@@ -15898,11 +16053,19 @@
}
Expr *SimpleExpr = RE->IgnoreParenCasts();
+ Expr::isModifiableLvalueResult IsLV =
+ VE->isModifiableLvalue(SemaRef.Context, &ELoc);
- if (!RE->IgnoreParenImpCasts()->isLValue()) {
- SemaRef.Diag(ELoc,
- diag::err_omp_expected_named_var_member_or_array_expression)
- << RE->getSourceRange();
+ if (!RE->IgnoreParenImpCasts()->isLValue() &&
+ IsLV != Expr::MLV_LValueCast) {
+ if (SemaRef.getLangOpts().OpenMP < 50) {
+ SemaRef.Diag(
+ ELoc, diag::err_omp_expected_named_var_member_or_array_expression)
+ << RE->getSourceRange();
+ } else {
+ SemaRef.Diag(ELoc, diag::err_omp_non_lvalue_in_map_or_motion_clauses)
+ << RE->getSourceRange();
+ }
continue;
}
@@ -15937,6 +16100,15 @@
MVLI.VarBaseDeclarations.push_back(nullptr);
continue;
}
+ if (const auto *BO = dyn_cast<BinaryOperator>(BE)) {
+ MVLI.UDMapperList.push_back(nullptr);
+ MVLI.ProcessedVarList.push_back(RE);
+ MVLI.VarComponents.resize(MVLI.VarComponents.size() + 1);
+ MVLI.VarComponents.back().append(CurComponents.begin(),
+ CurComponents.end());
+ MVLI.VarBaseDeclarations.push_back(nullptr);
+ continue;
+ }
// For the following checks, we rely on the base declaration which is
// expected to be associated with the last component. The declaration is
Index: clang/lib/CodeGen/CGOpenMPRuntime.cpp
===================================================================
--- clang/lib/CodeGen/CGOpenMPRuntime.cpp
+++ clang/lib/CodeGen/CGOpenMPRuntime.cpp
@@ -7406,6 +7406,7 @@
// S1 s;
// double *p;
// struct S2 *ps;
+ // struct S2 **ps_ptr;
// }
// S2 s;
// S2 *ps;
@@ -7542,6 +7543,48 @@
// &(ps->p), &(ps->p[0]), 33*sizeof(double), MEMBER_OF(4) | PTR_AND_OBJ | TO
// (*) the struct this entry pertains to is the 4th element in the list
// of arguments, hence MEMBER_OF(4)
+ //
+ // map(*p)
+ // &p, &(p[0]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map(*(p+3))
+ // &p, &(p[3]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map((p+3)+7)
+ // &p, &(p[10]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map(*(ps->p))
+ // ps, &(ps->p[0]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map(*(ps->p+3))
+ // ps, &(ps->p[3]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map((ps->p+3)[1])
+ // ps, &(ps->p[4]), 1*sizeof(float), TARGET_PARAM | TO | FROM
+ //
+ // map(*(ps->ps->p))
+ // ps, &(ps->ps), sizeof(S2*), TARGET_PARAM
+ // ps, &(ps->ps), sizeof(S2*), MEMBER_OF(1)
+ // &(ps->ps), &(ps->ps->p[0]), sizeof(double), MEMBER_OF(1) | PTR_AND_OBJ |
+ // FROM
+ //
+ // map((ps->ps->p)[3])
+ // ps, &(ps->ps), sizeof(S2*), TARGET_PARAM
+ // ps, &(ps->ps), sizeof(S2*), MEMBER_OF(1)
+ // &(ps->ps), &(ps->ps->p[3]), sizeof(double), MEMBER_OF(1) | PTR_AND_OBJ |
+ // FROM
+ //
+ // map(*(ps->ps->p+3))
+ // ps, &(ps->ps), sizeof(S2*), TARGET_PARAM
+ // ps, &(ps->ps), sizeof(S2*), MEMBER_OF(1)
+ // &(ps->ps), &(ps->ps->p[3]), sizeof(double), MEMBER_OF(1) | PTR_AND_OBJ |
+ // FROM
+ //
+ // map(*(ps->ps->ps_ptr))
+ // ps, &(ps->ps), sizeof(S2*), TARGET_PARAM
+ // ps, &(ps->ps), sizeof(S2*), MEMBER_OF(1)
+ // &(ps->ps_ptr), &(ps->ps->ps_ptr[0]), sizeof(S2*), MEMBER_OF(1) |
+ // PTR_AND_OBJ | TO | FROM
// Track if the map information being generated is the first for a capture.
bool IsCaptureFirstInfo = IsFirstComponentList;
@@ -7662,10 +7705,7 @@
if (Next == CE || IsPointer || IsFinalArraySection) {
// If this is not the last component, we expect the pointer to be
// associated with an array expression or member expression.
- assert((Next == CE ||
- isa<MemberExpr>(Next->getAssociatedExpression()) ||
- isa<ArraySubscriptExpr>(Next->getAssociatedExpression()) ||
- isa<OMPArraySectionExpr>(Next->getAssociatedExpression())) &&
+ assert((Next == CE || Next->getAssociatedExpression()->isLValue()) &&
"Unexpected expression");
Address LB = CGF.EmitOMPSharedLValue(I->getAssociatedExpression())
@@ -7781,6 +7821,8 @@
// mapped member. If the parent is "*this", then the value declaration
// is nullptr.
if (EncounteredME) {
+ assert(!isa<FunctionDecl>(EncounteredME->getMemberDecl()) &&
+ "Unexpected FunctionDecl");
const auto *FD = dyn_cast<FieldDecl>(EncounteredME->getMemberDecl());
unsigned FieldIndex = FD->getFieldIndex();
Index: clang/include/clang/Basic/DiagnosticSemaKinds.td
===================================================================
--- clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -9770,6 +9770,9 @@
def err_omp_lastprivate_conditional_non_scalar : Error<
"expected list item of scalar type in 'lastprivate' clause with 'conditional' modifier"
>;
+def err_omp_non_lvalue_in_map_or_motion_clauses: Error<
+ "expected lvalue with no function call in '#pragma omp target update' and '#pragma omp target map'"
+ >;
} // end of OpenMP category
let CategoryName = "Related Result Type Issue" in {
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
