[OpenMP5.0] Support more kinds of lvalues in map clauses

[jacobdweightman]

Summary

This patch aims to fully support this change from OpenMP 4.5 to 5.0:

The to and from clauses on the target update construct (see Section 2.12.6 on page 176), the depend clause on task generating constructs (see Section 2.17.11 on page 255), and the map clause (see Section 2.19.7.1 on page 315) were extended to allow any lvalue expression as a list item for C/C++.

Specifically, support is added for:

• calling functions which return references/constexprs
• cast expressions as part of a larger lvalue expression
• expressions with the ternary Elvis operators (A ? B : C and A ?: C)

This patch modifies Sema to accept such expressions. A few small changes in CodeGen were required to avoid asserts, but CodeGen appeared to already handle such expressions correctly.

[jacobdweightman]

clang/lib/Sema/SemaOpenMP.cpp

17141

It looks like this short-circuits and the false expression doesn't get visited if the true expression is "well-formed." We probably want to emit diagnostics and fail if either expression is bad (i.e. visit true expression && visit false expression), but that results in an incorrect component list.

I suspect this may indicate a deeper issue, since I'm not sure what the components should actually be. It seems like in general it should be a tree-like structure rather than a list-like structure, but that seems like it may require significant changes. Any thoughts on this or suggestions for improvement?

[jacobdweightman]

Ping

[ABataev]

In D91373#2408175, @jacobdweightman wrote:

Ping

1. Provide full diff, see https://www.llvm.org/docs/Phabricator.html for the guidance.
2. Need some more test with ast printing and codegen.
3. Would be good if you could run (at least locally) some tests to check that the changes really work.

[jacobdweightman]

Thanks for the initial feedback, @ABataev!

I've added one code gen test each for function calls, cast expressions, and ternaries. I haven't added any AST printing tests because I didn't think I changed anything about the AST, but please let me know what needs to be tested and I'll add it. Let me know if I need to be more thorough with the codegen tests as well!

I created this diff according to your instructions and now see that it includes the modified files in their entirety.

Prior to the first revision, I did run some test programs to verify that the changes work (these still work, since I haven't changed anything outside of tests for this revision). I had a test that looked like this for each of the three kinds of lvalue expressions this diff attempts to support:

#include <assert.h>

int main(void) {
    // test with map
    #pragma omp target data map(tofrom: <lvalue expression>)
    {
        <type of lvalue expression> xp = <lvalue expression>

        #pragma omp target
        <some mutation to lvalue expression>
    }

    assert(<mutation is reflected on host>);


    // test with update to/from
    int y;
    #pragma omp target data map(alloc: y) // we need a clause here, so do something useless
    {
        auto xp = <lvalue expression>;

        #pragma omp target update to(<lvalue expression>)

        #pragma omp target
        <some mutation to lvalue expression>

        #pragma omp target update from(<lvalue expression>)
        assert(<mutation is reflected on host>);
    }
}

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

16869

I don't think it should always return true. What about map(s.foo) where foo is a member function?

16939

Same, too general check, plus in some cases OVE->getSourceExpr() may return nullptr.

17065–17069

What is this for?

17110–17119

int a;
int &foo() {return a;}

...
#pragma omp target map(foo())
 foo() = 0;
...

How is this supposed to work?

17120–17139

Same questions here, how's the actual mapping is supposed to work? Need some more detailed description. I don't think this is going to be easy to implement it directly. We map the addresses of the base declarations but in your cases there is just no base declarations. SO, you need to create one. To me, this should look like this:

#pragma omp target map(<lvalue>)
{
  <lvalue> = xxx;
  yyy = <lvalue>;
}

V

auto &mapped = <lvalue>;
#pragma omp target map(mapped)
{
  mapped = xxx;
  yyy = mapped;
}

[jacobdweightman]

clang/lib/Sema/SemaOpenMP.cpp

16869

Hmm... I had previously added a test covering this on line 416 of target_map_messages.cpp which seemed to be passing, but not for the reason I thought. This program illustrates the difference:

struct Foo {
    int x;
    int &id() {
        return x;
    }
};

int x;
int &id() {
    return x;
}

int main(void) {
    Foo f;
    #pragma omp target map(tofrom: id, f.id)
    {}
}

The free function is parsed in bool Parser::ParseOpenMPVarList to this:

DeclRefExpr 0xfee3f8 'int &(void)' lvalue Function 0xfedc10 'id' 'int &(void)'

Whereas the method is parsed to this:

MemberExpr 0xfee438 '<bound member function type>' .id 0xfeda00
`-DeclRefExpr 0xfee418 'struct Foo' lvalue Var 0xfede98 'f' 'struct Foo'

Note that the former is an lvalue, whereas the latter is not. Therefore, the latter emits the error "early" inside of void checkMappableExpressionList due to the !RE->isLValue() check near SemaOpenMP.cpp:17668 rather than in the Visit* methods. I guess the current error message is misleading given that id is still an lvalue, though. Perhaps it would be good to add a new message specifically for free functions since they are lvalues, but that issue already exists in Clang today and feels out of scope for this change. If you disagree, I wouldn't mind adding it.

16939

I'm not exactly sure in what sense this check is too general, but perhaps it would be better to handle this together with the Elvis operator. For instance, I could have a VisitBinaryConditionalOperator which would "unwrap" the OpaqueValueExpression (OVA) directly rather than calling through to this method, since this is the only context in which we expect to handle OVAs. Let me know if this doesn't fully address your concern, though.

However, this usage of getSourceExpr appears to be consistent with other uses of OVAs that I see in Clang (I see a few similar examples in the static analyzer), but you make a good point. The Elvis operator's OVA should always have a source expression which refers to the condition without the implicit cast to bool. I'll add an assert that it isn't nullptr for now, but let me know if you have any other ideas.

17065–17069

This is a bit of a hack and definitely unclear to the reader. The Components.empty() check was added because an rvalue or a unary operator other than the de-referencing operator should be allowed in a sub-expression of the map clause list item, so long as the complete expression is an lvalue. As a minimal example, consider something like map(*(&x)). More usefully, one may cast pointer types so that a variable is mapped as a different type like map(*((int *) &x)). Without this check, the new casting tests which do this would fail.

A few ideas to make this more readable would be to use a temporary variable before the if statement like bool isFullExpression = Components.empty(); or add a comment explaining the above. Do you have any other ideas for how to improve this?

17110–17119

From my understanding of the spec, foo should be implicitly declared for both the host and the target. However, the user would be responsible for explicitly declaring a for the target if it isn't referenced in the target region. This test program seems to behave as I expect, with the result that a = 2:

#include <stdio.h>

int a;
#pragma omp declare target to(a)

int &foo() { return a; }

int main(void) {
    a = 1;

    #pragma omp target map(foo())
    foo() = 2;

    #pragma omp target update from(foo())

    printf("a = %d\n", a);
}

17120–17139

I think there's an issue with that source-level transformation when evaluating <lvalue> has side effects, since they would be performed 3 times in the first program but only once in the second program.

I might be off base here, but would there also be an issue if <lvalue> might share original storage with another list item? For example, shouldn't we issue an error when compiling the following program?

int a[10], b[10];
bool c;
#pragma omp target map(a, c ? a[1] : b[1])
{ ... }

If I'm missing something here, how would we go about implementing a transformation like that? Is it something that belongs in codegen? I think I need a bit more guidance here.

[jacobdweightman]

I separated the handling of ConditionalOperators and BinaryConditionalOperators in order to eliminate handling of OpaqueValueExpressions in general, and asserted that the OVE's SourceExpr is not null. I also made some minor changes to improve readability.

[jacobdweightman]

Ping

[ABataev]

Restore original formatting in test cases that were not directly affected by the patch. Also, I would start with a single kind of expression rather than trying to cover as many kinds of expressions as possible. It makes it easier to understand and to review it.

clang/lib/Sema/SemaOpenMP.cpp

16869

The mapping of functions is meaningless here, if they should return references to the globals and these globals should be marked as declare target

17065–17069

I would say that something like map(*((int *) &x)) should not be allowed. Something like map(*p) - yes, but not something that changes the original type.

17110–17119

Just like I said before, then mapping of the function is absolutely meaningless, we can just ignore it.

17120–17139

I assume side effects are not allowed throughout the standard though it is not explicitly expressed for the map clauses.
Plus, the fact that something is not allowed does not mean that the compiler should diagnose it. In general, it is just impossible, just in some cases.

17150–17151

&&?

17343

NextComponentExpr

17346

ShouldCheckForOverlap

17350

What if MemberExpr is a base of array section, subscript etc.?

clang/test/OpenMP/target_map_messages.cpp

95–114

Restore original formatting

191–208

Same here, remove unrelated changes

clang/lib/CodeGen/CGOpenMPRuntime.cpp

for (; I != CE; ++I) {
 
       // Get information on whether the element is a pointer. Have to do a
       // special treatment for array sections given that they are built-in
       // types.
       const auto *OASE =
           dyn_cast<OMPArraySectionExpr>(I->getAssociatedExpression());
       const auto *OAShE =
           dyn_cast<OMPArrayShapingExpr>(I->getAssociatedExpression());
-      const auto *UO = dyn_cast<UnaryOperator>(I->getAssociatedExpression());
-      const auto *BO = dyn_cast<BinaryOperator>(I->getAssociatedExpression());
       bool IsPointer =
           OAShE ||
           (OASE && OMPArraySectionExpr::getBaseOriginalType(OASE)
                        .getCanonicalType()
                        ->isAnyPointerType()) ||
           I->getAssociatedExpression()->getType()->isAnyPointerType();
-      bool IsNonDerefPointer = IsPointer && !UO && !BO && !IsNonContiguous;
+      bool IsNonDerefPointer =
+          IsPointer && !isa<UnaryOperator>(I->getAssociatedExpression()) &&
+          !isa<BinaryOperator>(I->getAssociatedExpression()) &&
+          !isa<CastExpr>(I->getAssociatedExpression()) &&
+          !isa<AbstractConditionalOperator>(I->getAssociatedExpression()) &&
+          !IsNonContiguous;
 
       if (OASE)
         ++DimSize;
 
       if (Next == CE || IsNonDerefPointer || 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()) ||
                 isa<OMPArrayShapingExpr>(Next->getAssociatedExpression()) ||
                 isa<UnaryOperator>(Next->getAssociatedExpression()) ||
-                isa<BinaryOperator>(Next->getAssociatedExpression())) &&
+                isa<BinaryOperator>(Next->getAssociatedExpression()) ||
+                isa<AbstractConditionalOperator>(
+                    Next->getAssociatedExpression())) &&
                "Unexpected expression");
 
         Address LB = Address::invalid();
         if (OAShE) {
           LB = Address(CGF.EmitScalarExpr(OAShE->getBase()),
                        CGF.getContext().getTypeAlignInChars(
                            OAShE->getBase()->getType()));
         } else {

clang/lib/Sema/SemaOpenMP.cpp

void emitErrorMsg() {
     } else {
       SemaRef.Diag(ELoc, diag::err_omp_non_lvalue_in_map_or_motion_clauses)
           << getOpenMPClauseName(CKind) << ERange;
     }
   }
 
 public:
   bool VisitDeclRefExpr(DeclRefExpr *DRE) {
-    if (!isa<VarDecl>(DRE->getDecl())) {
+    // Allow variables, and also allow functions if they are part of a call
+    // expression. However, we must not map functions themselves. This is
+    // prevented since a call expression will always be added to Components
+    // first.
+    if (!isa<VarDecl>(DRE->getDecl()) &&
+        (Components.empty() || !isa<FunctionDecl>(DRE->getDecl()))) {
       emitErrorMsg();
       return false;
     }
     assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
     RelevantExpr = DRE;
     // Record the component.
     Components.emplace_back(DRE, DRE->getDecl(), IsNonContiguous);
     return true;
   }
 
   bool VisitMemberExpr(MemberExpr *ME) {
     Expr *E = ME;
     Expr *BaseE = ME->getBase()->IgnoreParenCasts();
 
     if (isa<CXXThisExpr>(BaseE)) {
       assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
       // We found a base expression: this->Val.
       RelevantExpr = ME;
     } else {
       E = BaseE;
     }
 
+    // Allow results of method calls to be mapped.
+    if (isa<CXXMethodDecl>(ME->getMemberDecl())) {
+      RelevantExpr = ME;
+      Components.emplace_back(ME, ME->getMemberDecl(), IsNonContiguous);
+      return true;
+    }
+
     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;
bool VisitMemberExpr(MemberExpr *ME) {
     Components.emplace_back(ME, FD, IsNonContiguous);
     return RelevantExpr || Visit(E);
   }
 
   bool VisitArraySubscriptExpr(ArraySubscriptExpr *AE) {
     Expr *E = AE->getBase()->IgnoreParenImpCasts();
 
     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 RelevantExpr || Visit(E);
     }
 
     // If we got an array subscript that express the whole dimension we
bool VisitOMPArrayShapingExpr(OMPArrayShapingExpr *E) {
 
     // Record the component - we don't have any declaration associated.
     Components.emplace_back(E, nullptr, IsNonContiguous);
 
     return Visit(Base->IgnoreParenImpCasts());
   }
 
   bool VisitUnaryOperator(UnaryOperator *UO) {
-    if (SemaRef.getLangOpts().OpenMP < 50 || !UO->isLValue() ||
-        UO->getOpcode() != UO_Deref) {
+    bool isFullExpression = Components.empty();
+
+    if (SemaRef.getLangOpts().OpenMP < 50 ||
+        (isFullExpression &&
+         (!UO->isLValue() || UO->getOpcode() != UO_Deref))) {
       emitErrorMsg();
       return false;
     }
     if (!RelevantExpr) {
       // Record the component if haven't found base decl.
       Components.emplace_back(UO, nullptr, /*IsNonContiguous=*/false);
     }
     return RelevantExpr || Visit(UO->getSubExpr()->IgnoreParenImpCasts());
bool VisitCXXThisExpr(CXXThisExpr *CTE) {
     Components.emplace_back(CTE, nullptr, IsNonContiguous);
     return true;
   }
   bool VisitCXXOperatorCallExpr(CXXOperatorCallExpr *COCE) {
     assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
     Components.emplace_back(COCE, nullptr, IsNonContiguous);
     return true;
   }
+  bool VisitCallExpr(CallExpr *CE) {
+    if (SemaRef.getLangOpts().OpenMP < 50 ||
+        (Components.empty() && !CE->isLValue())) {
+      emitErrorMsg();
+      return false;
+    }
+    assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
+    Components.emplace_back(CE, nullptr, IsNonContiguous);
+    return RelevantExpr || Visit(CE->getCallee()->IgnoreParenImpCasts());
+  }
+  bool VisitCastExpr(CastExpr *CE) {
+    if (SemaRef.getLangOpts().OpenMP < 50 ||
+        (Components.empty() && !CE->isLValue())) {
+      emitErrorMsg();
+      return false;
+    }
+    assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
+    Components.emplace_back(CE, nullptr, false);
+    return RelevantExpr || Visit(CE->getSubExpr()->IgnoreParenImpCasts());
+  }
+  bool VisitConditionalOperator(ConditionalOperator *CO) {
+    if (SemaRef.getLangOpts().OpenMP < 50) {
+      emitErrorMsg();
+      return false;
+    }
+    assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
+    Components.emplace_back(CO, nullptr, false);
+    return RelevantExpr || Visit(CO->getTrueExpr()->IgnoreParenImpCasts()) ||
+           Visit(CO->getFalseExpr()->IgnoreParenImpCasts());
+  }
+  bool VisitBinaryConditionalOperator(BinaryConditionalOperator *BCO) {
+    if (SemaRef.getLangOpts().OpenMP < 50) {
+      emitErrorMsg();
+      return false;
+    }
+    assert(!RelevantExpr && "RelevantExpr is expected to be nullptr");
+    Components.emplace_back(BCO, nullptr, false);
+    Expr *TrueExpr =
+        dyn_cast<OpaqueValueExpr>(BCO->getTrueExpr())->getSourceExpr();
+    assert(TrueExpr && "Failed to resolve true expression of Elvis operator.");
+    return RelevantExpr || Visit(TrueExpr->IgnoreParenImpCasts()) ||
+           Visit(BCO->getFalseExpr()->IgnoreParenImpCasts());
+  }
   bool VisitStmt(Stmt *) {
     emitErrorMsg();
     return false;
   }
   const Expr *getFoundBase() const {
     return RelevantExpr;
   }
   explicit MapBaseChecker(
bool FoundError = DSAS->checkMappableExprComponentListsForDecl(
           }
           // If we find the same expression in the enclosing data environment,
           // that is legal.
           IsEnclosedByDataEnvironmentExpr = true;
           return false;
         }
 
         QualType DerivedType =
-            std::prev(CI)->getAssociatedDeclaration()->getType();
+            std::prev(CI)->getAssociatedExpression()->getType();
         SourceLocation DerivedLoc =
             std::prev(CI)->getAssociatedExpression()->getExprLoc();
 
         // 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.
         DerivedType = DerivedType.getNonReferenceType();
 
+        // There are restrictions on when arrays and structs may be mapped which
+        // are checked in this function. It is difficult to check these
+        // conditions at compile time for a general lvalue (e.g. for a ternary),
+        // so instead we whitelist the kind of expressions we would like to
+        // perform these checks on.
+        const Expr *nextComponentExpr = (CI == CE)
+                                            ? SI->getAssociatedExpression()
+                                            : CI->getAssociatedExpression();
+        const bool shouldCheckForOverlap =
+            isa<ArraySubscriptExpr>(nextComponentExpr) ||
+            isa<OMPArraySectionExpr>(nextComponentExpr) ||
+            isa<OMPArrayShapingExpr>(nextComponentExpr) ||
+            isa<MemberExpr>(nextComponentExpr);
+
         // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, C/C++, p.1]
         //  A variable for which the type is pointer and an array section
         //  derived from that variable must not appear as list items of map
         //  clauses of the same construct.
         //
         // Also, cover one of the cases in:
         // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.5]
         //  If any part of the original storage of a list item has corresponding
         //  storage in the device data environment, all of the original storage
         //  must have corresponding storage in the device data environment.
         //
-        if (DerivedType->isAnyPointerType()) {
+        if (shouldCheckForOverlap && DerivedType->isAnyPointerType()) {
           if (CI == CE || SI == SE) {
             SemaRef.Diag(
                 DerivedLoc,
                 diag::err_omp_pointer_mapped_along_with_derived_section)
                 << DerivedLoc;
             SemaRef.Diag(RE->getExprLoc(), diag::note_used_here)
                 << RE->getSourceRange();
             return true;
bool FoundError = DSAS->checkMappableExprComponentListsForDecl(
           }
         }
 
         // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.4]
         //  List items of map clauses in the same construct must not share
         //  original storage.
         //
         // An expression is a subset of the other.
-        if (CurrentRegionOnly && (CI == CE || SI == SE)) {
+        if (shouldCheckForOverlap && CurrentRegionOnly &&
+            (CI == CE || SI == SE)) {
           if (CKind == OMPC_map) {
             if (CI != CE || SI != SE) {
               // Allow constructs like this: map(s, s.ptr[0:1]), where s.ptr is
               // a pointer.
               auto Begin =
                   CI != CE ? CurComponents.begin() : StackComponents.begin();
               auto End = CI != CE ? CurComponents.end() : StackComponents.end();
               auto It = Begin;
bool FoundError = DSAS->checkMappableExprComponentListsForDecl(
           }
           SemaRef.Diag(RE->getExprLoc(), diag::note_used_here)
               << RE->getSourceRange();
           return true;
         }
 
         // The current expression uses the same base as other expression in the
         // data environment but does not contain it completely.
-        if (!CurrentRegionOnly && SI != SE)
+        if (shouldCheckForOverlap && !CurrentRegionOnly && SI != SE)
           EnclosingExpr = RE;
 
         // The current expression is a subset of the expression in the data
         // environment.
         IsEnclosedByDataEnvironmentExpr |=
             (!CurrentRegionOnly && CI != CE && SI == SE);
 
         return false;
for (Expr *RE : MVLI.VarList) {
     CurDeclaration = CurComponents.back().getAssociatedDeclaration();
     assert(CurDeclaration && "Null decl on map clause.");
     assert(
         CurDeclaration->isCanonicalDecl() &&
         "Expecting components to have associated only canonical declarations.");
 
     auto *VD = dyn_cast<VarDecl>(CurDeclaration);
     const auto *FD = dyn_cast<FieldDecl>(CurDeclaration);
+    const auto *FnD = dyn_cast<FunctionDecl>(CurDeclaration);
 
-    assert((VD || FD) && "Only variables or fields are expected here!");
+    assert((VD || FD || FnD) &&
+           "Only variables, fields, or functions are expected here!");
     (void)FD;
 
     // OpenMP 4.5 [2.15.5.1, map Clause, Restrictions, p.10]
     // threadprivate variables cannot appear in a map clause.
     // OpenMP 4.5 [2.10.5, target update Construct]
     // threadprivate variables cannot appear in a from clause.
     if (VD && DSAS->isThreadPrivate(VD)) {
       DSAStackTy::DSAVarData DVar = DSAS->getTopDSA(VD, /*FromParent=*/false);

clang/test/OpenMP/target_map_messages.cpp

-// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,ge50,ge51 -fopenmp -fopenmp-version=51 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 200 -x c %s -Wno-openmp -Wuninitialized
-
-// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp-simd -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp-simd -fopenmp-version=40 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp-simd -fopenmp-version=45 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp-simd -fopenmp-version=50 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -verify=expected,ge50,ge51 -fopenmp-simd -fopenmp-version=51 -ferror-limit 200 %s -Wno-openmp-target -Wuninitialized
-// RUN: %clang_cc1 -DCCODE -verify -fopenmp-simd -ferror-limit 200 -x c %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp -fopenmp-version=40 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp -fopenmp-version=45 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp -fopenmp-version=50 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,ge50,ge51 -fopenmp -fopenmp-version=51 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -DCCODE -verify -fopenmp -ferror-limit 400 -x c %s -Wno-openmp -Wuninitialized
+
+// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp-simd -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp-simd -fopenmp-version=40 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,lt50,lt51 -fopenmp-simd -fopenmp-version=45 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,ge50,lt51 -fopenmp-simd -fopenmp-version=50 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,ge50,ge51 -fopenmp-simd -fopenmp-version=51 -ferror-limit 400 %s -Wno-openmp-target -Wuninitialized
+// RUN: %clang_cc1 -DCCODE -verify -fopenmp-simd -ferror-limit 400 -x c %s -Wno-openmp-mapping -Wuninitialized
 #ifdef CCODE
 void foo(int arg) {
   const int n = 0;
 
   double marr[10][10][10];
 
   #pragma omp target map(marr[2][0:2][0:2]) // expected-error {{array section does not specify contiguous storage}}
   {}

 
 
 struct SREF {
   int &a;
   int b;
   SREF(int &a) : a(a) {}
 };
 
+int &id(int &x) {
+  return x;
+}
+
+int *id2(int *x) {
+  return x;
+}
+
 template <typename T, int I>
 struct SA {
   static int ss;
   #pragma omp threadprivate(ss) // expected-note {{defined as threadprivate or thread local}}
   float a;
   int b[12];
   float *c;
   T d;
void func(int arg) {
     #pragma omp target map(arg,a*2) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
     {}
     #pragma omp target map(arg,(c+1)[2]) // lt50-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}}
     {}
     #pragma omp target map(arg,a,d[:2]) // expected-error {{subscripted value is not an array or pointer}}
     {}
+#pragma omp target map(id)              // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+    {}
+#pragma omp target map(id(b[3]))        // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
+#pragma omp target map(id(id(b[3])))    // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
+#pragma omp target map(id2(&b[3]))      // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+    {}
+#pragma omp target map(*id2(&b[3]))     // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
 
-    #pragma omp target map(to:ss) // expected-error {{threadprivate variables are not allowed in 'map' clause}}
+#pragma omp target map(to \
+                       : ss) // expected-error {{threadprivate variables are not allowed in 'map' clause}}
     {}
 
-    #pragma omp target map(to:b,e)
+#pragma omp target map(to \
+                       : b, e)
     {}
-    #pragma omp target map(to:b,e) map(to:b) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
+#pragma omp target map(to: b, e) map(to: b) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
     {}
-    #pragma omp target map(to:b[:2],e)
+#pragma omp target map(to \
+                       : b[:2], e)
     {}
-    #pragma omp target map(to:b,e[:])
+#pragma omp target map(to \
+                       : b, e[:])
     {}
-    #pragma omp target map(b[-1:]) // expected-error {{array section must be a subset of the original array}}
+#pragma omp target map(b [-1:]) // expected-error {{array section must be a subset of the original array}}
     {}
-    #pragma omp target map(b[:-1]) // expected-error {{section length is evaluated to a negative value -1}}
+#pragma omp target map(b[:-1])  // expected-error {{section length is evaluated to a negative value -1}}
     {}
-    #pragma omp target map(b[true:true])
+#pragma omp target map(b [true:true])
     {}
 
     #pragma omp target map(: c,f) // expected-error {{missing map type}}
     {}
     #pragma omp target map(always, tofrom: c,f)
     {}
     #pragma omp target map(always, tofrom: c[1:2],f)
     {}
#pragma omp target map(b [true:true])
     // lt51-error@+2 2 {{incorrect map type modifier, expected 'always', 'close', or 'mapper'}}
     // expected-error@+1 {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
     #pragma omp target map( d, f, bf: a)
     {}
     // expected-error@+4 {{missing map type modifier}}
     // ge51-error@+3 {{incorrect map type modifier, expected 'always', 'close', 'mapper', or 'present'}}
     // lt51-error@+2 {{incorrect map type modifier, expected 'always', 'close', or 'mapper'}}
     // expected-error@+1 {{missing map type}}
-    #pragma omp target map( , f, : a)
+    #pragma omp target map(, f, : a)
     {}
-    #pragma omp target map(always close: a)   // expected-error {{missing map type}}
+    #pragma omp target map(always close: a) // expected-error {{missing map type}}
     {}
-    #pragma omp target map(always close bf: a)   // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+    #pragma omp target map(always close bf: a) // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
     {}
     // ge51-error@+3 {{incorrect map type modifier, expected 'always', 'close', 'mapper', or 'present'}}
     // lt51-error@+2 {{incorrect map type modifier, expected 'always', 'close', or 'mapper'}}
     // expected-error@+1 {{missing map type}}
     #pragma omp target map(always tofrom close: a)
     {}
     // ge51-error@+2 {{incorrect map type modifier, expected 'always', 'close', 'mapper', or 'present'}}
     // lt51-error@+1 {{incorrect map type modifier, expected 'always', 'close', or 'mapper'}}
     #pragma omp target map(tofrom from: a)
     {}
-    #pragma omp target map(close bf: a)   // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
+    #pragma omp target map(close bf: a)           // expected-error {{incorrect map type, expected one of 'to', 'from', 'tofrom', 'alloc', 'release', or 'delete'}}
     {}
-    #pragma omp target map(([b[I]][bf])f)  // lt50-error {{expected ',' or ']' in lambda capture list}} lt50-error {{expected ')'}} lt50-note {{to match this '('}}
+    #pragma omp target map(([b[I]][bf])f) // lt50-error {{expected ',' or ']' in lambda capture list}} lt50-error {{expected ')'}} lt50-note {{to match this '('}}
     {}
     return;
   }
 };
 
 struct SB {
   unsigned A;
   unsigned B;
   float Arr[100];
   float *Ptr;
   float *foo() {
     return &Arr[0];
   }
+  float &bar() {
+    return Arr[0];
+  }
 };
 
 struct SC {
   unsigned A : 2;
   unsigned B : 3;
   unsigned C;
   unsigned D;
   float Arr[100];

 #pragma omp target map(r.RPtrS[0], r.RPtrS->B) // expected-error {{same pointer dereferenced in multiple different ways in map clause expressions}} expected-note {{used here}}
   {}
 #pragma omp target map(r.RPtrS, r.RPtrS->B)    // expected-error {{pointer cannot be mapped along with a section derived from itself}} expected-note {{used here}}
   {}
 #pragma omp target map(r.RPtrS->A, r.RPtrS->B)
   {}
 #pragma omp target map(r.S.Arr[:12])
   {}
-#pragma omp target map(r.S.foo() [:12]) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+#pragma omp target map(r.S.foo)         // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(r.S.foo())       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(r.S.foo() [:12]) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target map(r.S.bar())       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
   {}
 #pragma omp target map(r.C, r.D)
   {}
 #pragma omp target map(r.C, r.C)     // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
   {}
 #pragma omp target map(r.C) map(r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
   {}
 #pragma omp target map(r.C, r.S)     // this would be an error only caught at runtime - Sema would have to make sure there is not way for the missing data between fields to be mapped somewhere else.
#pragma omp target map(to: x)
   foo();
 #pragma omp target map(to: to)
   foo();
 #pragma omp target map(to)
   foo();
 #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) // lt50-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error 2 {{expected addressable lvalue in 'map' clause}}
+#pragma omp target data map(tofrom \
+                            : argc > 0 ? x : y) // lt50-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom \
+                            : x ?: y) // lt50-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}} warn-warning 2 {{Type 'const S2' is not trivially copyable and not guaranteed to be mapped correctly}} warn-warning 2 {{Type 'const S3' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(ba) // warn-warning 2 {{Type 'const S2 [5]' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(ca) // warn-warning 2 {{Type 'const S3 [5]' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(da)
 #pragma omp target data map(S2::S2s)
 #pragma omp target data map(S2::S2sc)
public:
   virtual void foo() = 0;
 } *s8;
 
 class S9 {
 public:
   virtual void foo() {}
 } s9;
 
+class S10 : public S9 {
+public:
+  int a;
+  void foo() {}
+} s10;
 
 int main(int argc, char **argv) {
   const int d = 5;
   const int da[5] = { 0 };
   S4 e(4);
   S5 g(5);
   int i;
   int &j = i;
#pragma omp target map(to: x)
   foo();
 #pragma omp target map(to: to)
   foo();
 #pragma omp target map(to)
   foo();
 #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]) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+#pragma omp target data map(tofrom \
+                            : argc > 0 ? argv[1] : argv[2]) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom \
+                            : argv[1] ?: argv[2]) // lt50-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}} warn-warning {{Type 'const S2' is not trivially copyable and not guaranteed to be mapped correctly}} warn-warning {{Type 'const S3' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(argv[1])
 #pragma omp target data map(ba) // warn-warning {{Type 'const S2 [5]' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(ca) // warn-warning {{Type 'const S3 [5]' is not trivially copyable and not guaranteed to be mapped correctly}}
 #pragma omp target data map(da)
 #pragma omp target data map(S2::S2s)

 #pragma omp target map(**(*offset+BB+*a)) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
   {}
 #pragma omp target map(**(*(*(&offset))+BB+*a)) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
   {}
 #pragma omp target map(*(a+(a))) // expected-error {{invalid operands to binary expression ('int *' and 'int *')}}
   {}
 #pragma omp target map(*(1+*a+*a)) // expected-error {{indirection requires pointer operand ('int' invalid)}}
   {}
+#pragma omp target map(reinterpret_cast <int *>(&s10))  // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(*reinterpret_cast <int *>(&s10)) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target map(static_cast <S9>(s10))           // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(*static_cast <S9 *>(&s10))       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target map(*dynamic_cast <S9 *>(&s10))      // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target map(const_cast <S10 *>(&s10))        // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(*const_cast <S10 *>(&s10))       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target map((S9)s10)                         // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target map(*((S9 *)&s10))                   // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
 
 #pragma omp target map(delete: a) // expected-error {{map type 'delete' is not allowed for '#pragma omp target'}}
   {}
 #pragma omp target map(release: a) // expected-error {{map type 'release' is not allowed for '#pragma omp target'}}
   {}
 
   int marr[10][10], iarr[5];
 

clang/test/OpenMP/target_parallel_for_map_messages.cpp

 #pragma omp target parallel for map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel for map(tofrom \
                                     : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel for map(tofrom \
+                                    : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target parallel for map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(ba)
   for (i = 0; i < argc; ++i) foo();
#pragma omp target parallel for map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i)
     foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel for map(tofrom \
                                     : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel for map(tofrom \
+                                    : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target parallel for map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_parallel_for_simd_map_messages.cpp

 #pragma omp target parallel for simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel for simd map(tofrom \
                                          : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel for simd map(tofrom \
+                                         : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target parallel for simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(ba)
   for (i = 0; i < argc; ++i) foo();
#pragma omp target parallel for simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to, z)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i)
     foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel for simd map(tofrom \
                                          : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel for simd map(tofrom \
+                                         : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target parallel for simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target parallel for simd map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_parallel_map_messages.cpp

 #pragma omp target parallel map(to: to)
   foo();
 #pragma omp target parallel map(to)
   foo();
 #pragma omp target parallel map(to, x)
   foo();
 #pragma omp target parallel map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel map(tofrom \
                                 : argc > 0 ? x : y)
   foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel map(tofrom \
+                                : x ?: y)
+  foo();
 #pragma omp target parallel map(argc)
   foo();
 #pragma omp target parallel map(S1) // expected-error {{'S1' does not refer to a value}}
   foo();
 #pragma omp target parallel map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   foo();
 #pragma omp target parallel map(ba)
   foo();

 #pragma omp target parallel map(to: to)
   foo();
 #pragma omp target parallel map(to)
   foo();
 #pragma omp target parallel map(to, x)
   foo();
 #pragma omp target parallel map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target parallel map(tofrom \
                                 : argc > 0 ? argv[1] : argv[2])
   foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target parallel map(tofrom \
+                                : argv[1] ?: argv[2])
+  foo();
 #pragma omp target parallel map(argc)
   foo();
 #pragma omp target parallel map(S1) // expected-error {{'S1' does not refer to a value}}
   foo();
 #pragma omp target parallel map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   foo();
 #pragma omp target parallel map(argv[1])
   foo();

clang/test/OpenMP/target_simd_map_messages.cpp

 #pragma omp target simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target simd map(tofrom \
                             : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target simd map(tofrom \
+                            : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(ba)
   for (i = 0; i < argc; ++i) foo();
#pragma omp target simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i)
     foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target simd map(tofrom \
                             : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target simd map(tofrom \
+                            : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target simd map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_teams_distribute_map_messages.cpp

 #pragma omp target teams distribute map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute map(tofrom \
                                         : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute map(tofrom \
+                                        : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(ba)
   for (i = 0; i < argc; ++i) foo();
#pragma omp target teams distribute map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i)
     foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute map(tofrom \
                                         : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute map(tofrom \
+                                        : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_teams_distribute_parallel_for_map_messages.cpp

 #pragma omp target teams distribute parallel for map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute parallel for map(tofrom \
                                                      : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute parallel for map(tofrom \
+                                                     : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute parallel for map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(ba)
   for (i = 0; i < argc; ++i) foo();

 #pragma omp target teams distribute parallel for map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute parallel for map(tofrom \
                                                      : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute parallel for map(tofrom \
+                                                     : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute parallel for map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_teams_distribute_parallel_for_simd_map_messages.cpp

 #pragma omp target teams distribute parallel for simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute parallel for simd map(tofrom \
                                                           : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute parallel for simd map(tofrom \
+                                                          : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute parallel for simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(ba)
   for (i = 0; i < argc; ++i) foo();

 #pragma omp target teams distribute parallel for simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to, x, z)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute parallel for simd map(tofrom \
                                                           : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute parallel for simd map(tofrom \
+                                                          : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute parallel for simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute parallel for simd map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_teams_distribute_simd_map_messages.cpp

 #pragma omp target teams distribute simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i) foo();
-// ge50-error@+3 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute simd map(tofrom \
                                              : argc > 0 ? x : y)
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute simd map(tofrom \
+                                             : x ?: y)
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(ba)
   for (i = 0; i < argc; ++i) foo();
#pragma omp target teams distribute simd map(to: to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to, x)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
   for (i = 0; i < argc; ++i)
     foo();
-// ge50-error@+3 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams distribute simd map(tofrom \
                                              : argc > 0 ? argv[1] : argv[2])
   for (i = 0; i < argc; ++i) foo();
+// lt50-error@+2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target teams distribute simd map(tofrom \
+                                             : argv[1] ?: argv[2])
+  for (i = 0; i < argc; ++i)
+    foo();
 #pragma omp target teams distribute simd map(argc)
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(S1) // expected-error {{'S1' does not refer to a value}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(a, b, c, d, f) // expected-error {{incomplete type 'S1' where a complete type is required}}
   for (i = 0; i < argc; ++i) foo();
 #pragma omp target teams distribute simd map(argv[1])
   for (i = 0; i < argc; ++i) foo();

clang/test/OpenMP/target_teams_map_messages.cpp

 
 void xxx(int argc) {
   int map; // expected-note {{initialize the variable 'map' to silence this warning}}
 #pragma omp target teams map(tofrom: map) // expected-warning {{variable 'map' is uninitialized when used here}}
   for (int i = 0; i < 10; ++i)
     ;
 }
 
+int &id(int &x) {
+  return x;
+}
+
+int *id2(int *x) {
+  return x;
+}
+
 template <typename T, int I>
 struct SA {
   static int ss;
   #pragma omp threadprivate(ss) // expected-note {{defined as threadprivate or thread local}}
   float a;
   int b[12];
   float *c;
   T d;
void func(int arg) {
     {}
     // lt50-error@+1 {{expected expression containing only member accesses and/or array sections based on named variables}}
     #pragma omp target teams map(arg,(c+1)[2])
     {}
     #pragma omp target teams map(arg,a[:2],d) // expected-error {{subscripted value is not an array or pointer}}
     {}
     #pragma omp target teams map(arg,a,d[:2]) // expected-error {{subscripted value is not an array or pointer}}
     {}
+    #pragma omp target teams map(id) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+    {}
+    #pragma omp target teams map(id(b[3])) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
+    #pragma omp target teams map(id(id(b[3]))) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
+    #pragma omp target teams map(id2(&b[3])) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+    {}
+    #pragma omp target teams map(*id2(&b[3])) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+    {}
 
     #pragma omp target teams map(to:ss) // expected-error {{threadprivate variables are not allowed in 'map' clause}}
     {}
 
     #pragma omp target teams map(to:b,e)
     {}
     #pragma omp target teams map(to:b,e) map(to:b) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
     {}

 struct SB {
   unsigned A;
   unsigned B;
   float Arr[100];
   float *Ptr;
   float *foo() {
     return &Arr[0];
   }
+  float &bar() {
+    return Arr[0];
+  }
 };
 
 struct SC {
   unsigned A : 2;
   unsigned B : 3;
   unsigned C;
   unsigned D;
   float Arr[100];
void SAclient(int arg) {
   #pragma omp target teams map(r.RPtrS[0], r.RPtrS->B) // expected-error {{same pointer dereferenced in multiple different ways in map clause expressions}} expected-note {{used here}}
   {}
   #pragma omp target teams map(r.RPtrS, r.RPtrS->B) // expected-error {{pointer cannot be mapped along with a section derived from itself}} expected-note {{used here}}
   {}
   #pragma omp target teams map(r.RPtrS->A, r.RPtrS->B)
   {}
   #pragma omp target teams map(r.S.Arr[:12])
   {}
-  // ge50-error@+2 {{expected addressable lvalue in 'map' clause}}
+  #pragma omp target teams map(r.S.foo) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+  #pragma omp target teams map(r.S.foo()) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+  #pragma omp target teams map(*(r.S.foo())) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
   // lt50-error@+1 {{expected expression containing only member accesses and/or array sections based on named variables}}
   #pragma omp target teams map(r.S.foo()[:12])
   {}
+  #pragma omp target teams map(r.S.bar()) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
   #pragma omp target teams map(r.C, r.D)
   {}
   #pragma omp target teams map(r.C, r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
   {}
   #pragma omp target teams map(r.C) map(r.C) // expected-error {{variable already marked as mapped in current construct}} expected-note {{used here}}
   {}
   #pragma omp target teams map(r.C, r.S)  // this would be an error only caught at runtime - Sema would have to make sure there is not way for the missing data between fields to be mapped somewhere else.
   {}

 struct S6;
 
 template<>
 struct S6<int>
 {
    virtual void foo();
 };
 
+class S9 {
+public:
+  virtual void foo() {}
+} s9;
+
+class S10 : public S9 {
+  int a;
+  void foo(){};
+} s10;
+
 S3 h;
 #pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}}
 
 typedef int from;
 
 template <typename T, int I> // expected-note {{declared here}}
 T tmain(T argc) {
   const T d = 5;

 #pragma omp target teams map(to: to)
   foo();
 #pragma omp target teams map(to)
   foo();
 #pragma omp target teams map(to, x)
   foo();
 
 #pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-// ge50-error@+2 2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+1 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)
+// lt50-error@+1 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: x ?: y)
 
 #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}}
 #pragma omp target data map(ba)
 #pragma omp target data map(ca)
 #pragma omp target data map(da)
 #pragma omp target data map(S2::S2s)

 #pragma omp target teams map(to: to)
   foo();
 #pragma omp target teams map(to)
   foo();
 #pragma omp target teams map(to, x)
   foo();
 
 #pragma omp target data map(to x) // expected-error {{expected ',' or ')' in 'map' clause}}
-// ge50-error@+2 {{expected addressable lvalue in 'map' clause}}
 // lt50-error@+1 {{expected 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])
+// lt50-error@+1 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target data map(tofrom: argv[1] ?: argv[2])
 #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}}
 #pragma omp target data map(argv[1])
 #pragma omp target data map(ba)
 #pragma omp target data map(ca)
 #pragma omp target data map(da)
 #pragma omp target data map(S2::S2s)
#pragma omp target teams map(**(*offset+BB+*a))
   {}
 // lt50-error@+1 {{expected expression containing only member accesses and/or array sections based on named variables}}
 #pragma omp target teams map(**(*(*(&offset))+BB+*a))
   {}
 #pragma omp target teams map(*(a+(a))) // expected-error {{invalid operands to binary expression ('int *' and 'int *')}}
   {}
 #pragma omp target teams map(*(1+*a+*a)) // expected-error {{indirection requires pointer operand ('int' invalid)}}
   {}
+#pragma omp target teams map(reinterpret_cast <int *>(&s10))  // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target teams map(*reinterpret_cast <int *>(&s10)) // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target teams map(static_cast <S9>(s10))           // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target teams map(*static_cast <S9 *>(&s10))       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target teams map(*dynamic_cast <S9 *>(&s10))      // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target teams map(const_cast <S10 *>(&s10))        // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target teams map(*const_cast <S10 *>(&s10))       // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
+#pragma omp target teams map((S9)s10)                         // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}} ge50-error {{expected addressable lvalue in 'map' clause}}
+  {}
+#pragma omp target teams map(*((S9 *)&s10))                   // lt50-error {{expected expression containing only member accesses and/or array sections based on named variables}}
+  {}
 
 #pragma omp target teams map(delete: j) // expected-error {{map type 'delete' is not allowed for '#pragma omp target teams'}}
   {}
 #pragma omp target teams map(release: j) // expected-error {{map type 'release' is not allowed for '#pragma omp target teams'}}
   {}
 
   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}}
 }
 #endif

clang/test/OpenMP/target_update_codegen.cpp

void foo(int arg) {
   // CK25: store i8* [[PC1]], i8** [[PTRS_2]],
   // CK25-DAG: call void @__tgt_target_data_update_mapper(%struct.ident_t* @{{.+}}, i64 -1, i32 3, i8** [[GEPBP:%.+]], i8** [[GEPP:%.+]], {{.+}}getelementptr {{.+}}[3 x i{{.+}}]* [[MSIZE]], {{.+}}getelementptr {{.+}}[3 x i{{.+}}]* [[MTYPE]]{{.+}})
   // CK25-DAG: [[GEPBP]] = getelementptr inbounds {{.+}}[[BP]]
   // CK25-DAG: [[GEPP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
 
 #pragma omp target update to(arr[0:2][arg:][1:4], x, farr[0:2][1:2])
   { ++arg; }
 }
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK26 -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 CK26 --check-prefix CK26-64
+// RUN: %clang_cc1 -DCK26 -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 CK26 --check-prefix CK26-64
+// RUN: %clang_cc1 -DCK26 -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 CK26 --check-prefix CK26-32
+// RUN: %clang_cc1 -DCK26 -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 CK26 --check-prefix CK26-32
+
+// RUN: %clang_cc1 -DCK26 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK26 -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-ONLY19 %s
+// SIMD-ONLY19-NOT: {{__kmpc|__tgt}}
+#ifdef CK26
+
+// CK26: id
+int &id(int &x) {
+  return x;
+}
+
+// CK26: foo
+void foo(int arg) {
+  int x;
+
+  // CK26: %x = alloca i32, align {{4|8}}
+  // CK26: [[OFFLOAD_BP:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK26: [[OFFLOAD_PTRS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK26: [[OFFLOAD_MAPPERS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK26: [[RESULT:%.+]] = call {{.+}} i32* @_Z2idRi(i32* {{.+}} %x)
+  // CK26: [[R0:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK26: [[R1:%.+]] = bitcast i8** [[R0]] to i32* (i32*)**
+  // CK26: store i32* (i32*)* @_Z2idRi, i32* (i32*)** [[R1]], align {{4|8}}
+  // CK26: [[R2:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK26: [[R3:%.+]] = bitcast i8** [[R2]] to i32**
+  // CK26: store i32* [[RESULT]], i32** [[R3]], align {{4|8}}
+  // CK26: [[R4:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_MAPPERS]], {{i32|i64}} 0, {{i32|i64}} 0
+  // CK26: store i8* null, i8** [[R4]], align {{4|8}}
+  // CK26: [[GEPBP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK26: [[GEPP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK26-DAG: call void @__tgt_target_data_update_mapper(%struct.ident_t* @{{.+}}, i64 -1, i32 1, i8** [[GEPBP]], i8** [[GEPP]], {{.+}}getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes, i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes, i32 0, i32 0), i8** null, i8** null)
+
+#pragma omp target update to(id(x))
+}
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-64
+// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-64
+// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-32
+// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-32
+
+// RUN: %clang_cc1 -DCK27 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK27 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK27 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK27 -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-ONLY19 %s
+// SIMD-ONLY19-NOT: {{__kmpc|__tgt}}
+#ifdef CK27
+
+// CK27: wrapper
+struct wrapper {
+  int wrapped;
+};
+
+// CK27: foo
+void foo(int arg) {
+  wrapper x;
+  x.wrapped = 1;
+  wrapper *xp = &x;
+
+  // CK27: %x = alloca %struct.wrapper, align 4
+  // CK27: %xp = alloca %struct.wrapper*, align {{4|8}}
+  // CK27: [[OFFLOAD_BP:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK27: [[OFFLOAD_PTRS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK27: [[OFFLOAD_MAPPERS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK27: store i32 %arg, i32* %arg.addr, align 4
+  // CK27: %wrapped = getelementptr inbounds %struct.wrapper, %struct.wrapper* %x, i32 0, i32 0
+  // CK27: store i32 1, i32* %wrapped, align 4
+  // CK27: store %struct.wrapper* %x, %struct.wrapper** %xp, align {{4|8}}
+  // CK27: [[R0:%.+]] = load %struct.wrapper*, %struct.wrapper** %xp, align {{4|8}}
+  // CK27: [[R1:%.+]] = load %struct.wrapper*, %struct.wrapper** %xp, align {{4|8}}
+  // CK27: [[R2:%.+]] = bitcast %struct.wrapper* [[R1]] to i32*
+  // CK27: [[R3:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK27: [[R4:%.+]] = bitcast i8** [[R3]] to %struct.wrapper**
+  // CK27: store %struct.wrapper* [[R0]], %struct.wrapper** [[R4]], align {{4|8}}
+  // CK27: [[R5:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK27: [[R6:%.+]] = bitcast i8** [[R5]] to i32**
+  // CK27: store i32* [[R2]], i32** [[R6]], align {{4|8}}
+  // CK27: [[R7:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_MAPPERS]], {{i32|i64}} 0, {{i32|i64}} 0
+  // CK27: store i8* null, i8** [[R7]], align {{4|8}}
+  // CK27: [[GEPBP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK27: [[GEPP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK27: call void @__tgt_target_data_update_mapper(%struct.ident_t* @1, i64 -1, i32 1, i8** [[GEPBP]], i8** [[GEPP]], i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes, i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes, i32 0, i32 0), i8** null, i8** null)
+#pragma omp target update to(*reinterpret_cast <int *>(xp))
+}
+#endif
+///==========================================================================///
+// RUN: %clang_cc1 -DCK28 -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 CK28 --check-prefix CK28-64
+// RUN: %clang_cc1 -DCK28 -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 CK28 --check-prefix CK28-64
+// RUN: %clang_cc1 -DCK28 -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 CK28 --check-prefix CK28-32
+// RUN: %clang_cc1 -DCK28 -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 CK28 --check-prefix CK28-32
+
+// RUN: %clang_cc1 -DCK28 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK28 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK28 -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-ONLY19 %s
+// RUN: %clang_cc1 -DCK28 -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-ONLY19 %s
+// SIMD-ONLY19-NOT: {{__kmpc|__tgt}}
+#ifdef CK28
+
+// CK28: foo
+void foo(int arg) {
+  bool which = arg > 0;
+  int a = 1;
+  int b = 2;
+
+  // CK28: %arg.addr = alloca i32, align 4
+  // CK28: %which = alloca i8, align 1
+  // CK28: %a = alloca i32, align 4
+  // CK28: %b = alloca i32, align 4
+  // CK28: [[OFFLOAD_BP:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK28: [[OFFLOAD_PTRS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK28: [[OFFLOAD_MAPPERS:%.+]] = alloca [1 x i8*], align {{4|8}}
+  // CK28: store i32 %arg, i32* %arg.addr, align 4
+  // CK28: [[R0:%.+]] = load i32, i32* %arg.addr, align 4
+  // CK28: %cmp = icmp sgt i32 [[R0]], 0
+  // CK28: %frombool = zext i1 %cmp to i8
+  // CK28: store i8 %frombool, i8* %which, align 1
+  // CK28: store i32 1, i32* %a, align 4
+  // CK28: store i32 2, i32* %b, align 4
+  // CK28: [[R1:%.+]] = load i8, i8* %which, align 1
+  // CK28: %tobool = trunc i8 [[R1]] to i1
+  // CK28: br i1 %tobool, label %[[TRUE_BRANCH:.+]], label %[[FALSE_BRANCH:.+]]
+
+  // CK28: [[TRUE_BRANCH]]: ; preds = %entry
+  // CK28:   br label %[[COND_END:.+]]
+
+  // CK28: [[FALSE_BRANCH]]: ; preds = %entry
+  // CK28:   br label %[[COND_END]]
+
+  // CK28: [[COND_END]]: ; preds = %[[FALSE_BRANCH]], %[[TRUE_BRANCH]]
+  // CK28:   [[LVALUE:%.+]] = phi i32* [ %a, %[[TRUE_BRANCH]] ], [ %b, %[[FALSE_BRANCH]] ]
+  // CK28:   [[R2:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK28:   [[R3:%.+]] = bitcast i8** [[R2]] to i32**
+  // CK28:   store i32* %a, i32** [[R3]], align {{4|8}}
+  // CK28:   [[R4:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK28:   [[R5:%.+]] = bitcast i8** [[R4]] to i32**
+  // CK28:   store i32* [[LVALUE]], i32** [[R5]], align {{4|8}}
+  // CK28:   [[R6:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_MAPPERS]], {{i32|i64}} 0, {{i32|i64}} 0
+  // CK28:   store i8* null, i8** [[R6]], align {{4|8}}
+  // CK28:   [[GEPBP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP]], i32 0, i32 0
+  // CK28:   [[GEPP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_PTRS]], i32 0, i32 0
+  // CK28:   call void @__tgt_target_data_update_mapper(%struct.ident_t* @1, i64 -1, i32 1, i8** [[GEPBP]], i8** [[GEPP]], i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes, i32 0, i32 0), i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes, i32 0, i32 0), i8** null, i8** null)
+
+#pragma omp target update to(which ? a : b)
+}
 
 #endif
 #endif

clang/test/OpenMP/target_update_from_messages.cpp

-// RUN: %clang_cc1 -verify=expected,le50 -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=expected,le50 -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 -verify=expected,le50 -fopenmp-simd -ferror-limit 100 %s -Wno-openmp-mapping -Wuninitialized
+// RUN: %clang_cc1 -verify=expected,le50 -fopenmp-simd -ferror-limit 200 %s -Wno-openmp-mapping -Wuninitialized
 
 void foo() {
 }
 
 bool foobool(int argc) {
   return argc;
 }
 
+int &id(int &x) {
+  return x;
+}
+
+int *id2(int *x) {
+  return x;
+}
+
 struct S1; // expected-note 2 {{declared here}}
 extern S1 a;
 class S2 {
   mutable int a;
 public:
   S2():a(0) { }
   S2(S2 &s2):a(s2.a) { }
   static float S2s;

 #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'}}
 #pragma omp target update from(*(this+this)) // expected-error {{invalid operands to binary expression ('S8 *' and 'S8 *')}}
+
+#pragma omp target update from(id)               // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(id(this->a))      // 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(id(id(this->a)))  // 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(*id2(&(this->a))) // 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(id2(&(this->a)))  // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
   }
 };
 
 S3 h;
 #pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}}
 
 typedef int to;
 

 #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) // le50-error 2 {{expected addressable lvalue in 'from' clause}} 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) // le50-error 2 {{expected addressable lvalue in 'from' clause}} le45-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) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update from(x ?: y)           // 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)
 #pragma omp target update from(h) // expected-error {{threadprivate variables are not allowed in 'from' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(k), to(k) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update from(t), from(t[:5]) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update from(da)

 #pragma omp target data map(to: s7.i)
   {
 #pragma omp target update from(s7.x)
   }
 
   return 0;
 }
 
+class S9 {
+public:
+  virtual void foo() {}
+} s9;
+
+class S10 : public S9 {
+  int a;
+  void foo() {}
+} s10;
+
 int main(int argc, char **argv) {
   const int d = 5;
   const int da[5] = { 0 };
   S4 e(4);
   S5 g(5);
   int i, t[20];
   int &j = i;
   int *k = &j;

 #pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(x)
 #pragma omp target update from(t[:i])
 #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) // le50-error {{expected addressable lvalue in 'from' clause}} 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(argc > 0 ? x : 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(x ?: 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(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)
 #pragma omp target update from(h) // expected-error {{threadprivate variables are not allowed in 'from' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update from(k), to(k) // expected-error {{variable can appear only once in OpenMP 'target update' construct}} expected-note {{used here}}
 #pragma omp target update from(t), from(t[:5]) // expected-error {{variable can appear only once in OpenMP 'target update' construct}} expected-note {{used here}}
 #pragma omp target update from(da)
 #pragma omp target update from(da[:4])

 #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))) // 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+(m))) // expected-error {{invalid operands to binary expression ('int *' and 'int *')}}
 #pragma omp target update from(*(1+y+y)) // expected-error {{indirection requires pointer operand ('int' invalid)}}
 #pragma omp target data map(to: s7.i)
   {
 #pragma omp target update from(s7.x)
   }
+#pragma omp target update from(id)        // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(id(x))     // 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(id(id(x))) // 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(*id2(&x))  // 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(id2(&x))   // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+
+#pragma omp target update from(reinterpret_cast <int *>(&s10))  // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*reinterpret_cast <int *>(&s10)) // 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(static_cast <S9>(s10))           // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*static_cast <S9 *>(&s10))       // 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(*dynamic_cast <S9 *>(&s10))      // 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(const_cast <S10 *>(&s10))        // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*const_cast <S10 *>(&s10))       // 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((S9)s10)                         // 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'}} le50-error {{expected addressable lvalue in 'from' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update from(*((S9 *)&s10))                   // 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'}}
 
   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}}
 }
 

clang/test/OpenMP/target_update_to_messages.cpp

-// RUN: %clang_cc1 -verify=expected,le50 -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=expected,le50 -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
+// RUN: %clang_cc1 -verify=expected,le50 -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 -verify=expected,le50 -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
 
 void foo() {
 }
 
 bool foobool(int argc) {
   return argc;
 }
 
+int &id(int &x) {
+  return x;
+}
+
+int *id2(int *x) {
+  return x;
+}
+
 struct S1; // expected-note 2 {{declared here}}
 extern S1 a;
 class S2 {
   mutable int a;
 public:
   S2():a(0) { }
   S2(S2 &s2):a(s2.a) { }
   static float S2s;

 #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'}}
 #pragma omp target update to(*(this+this)) // expected-error {{invalid operands to binary expression ('S8 *' and 'S8 *')}}
     {}
 
     double marr[10][5][10];
 #pragma omp target update to(marr [0:1][2:4][1:2]) // le45-error {{array section does not specify contiguous storage}} le45-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
     {}
+
+#pragma omp target update to(id)               // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(id(this->a))      // 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(id(id(this->a)))  // 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(*id2(&(this->a))) // 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(id2(&(this->a)))  // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
   }
 };
 
 S3 h;
 #pragma omp threadprivate(h) // expected-note 2 {{defined as threadprivate or thread local}}
 
 typedef int from;
 

 #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) // le50-error 2 {{expected addressable lvalue in 'to' clause}} 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) // le50-error 2 {{expected addressable lvalue in 'to' clause}} le45-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) // le45-error 2 {{expected expression containing only member accesses and/or array sections based on named variables}}
+#pragma omp target update to(x ?: y)           // 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)
 #pragma omp target update to(h) // expected-error {{threadprivate variables are not allowed in 'to' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(k), from(k) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update to(t), to(t[:5]) // expected-error 2 {{variable can appear only once in OpenMP 'target update' construct}} expected-note 2 {{used here}}
 #pragma omp target update to(da)
 #pragma omp target update to(da[:4])

 #pragma omp target update to(x, s7.p[:]) // expected-error {{section length is unspecified and cannot be inferred because subscripted value is not an array}}
 #pragma omp target data map(to: s7.i)
   {
 #pragma omp target update to(s7.x)
   }
   return 0;
 }
 
+class S9 {
+public:
+  virtual void foo() {}
+} s9;
+
+class S10 : public S9 {
+  int a;
+  void foo() {}
+} s10;
+
 int main(int argc, char **argv) {
   const int d = 5;
   const int da[5] = { 0 };
   S4 e(4);
   S5 g(5);
   int i, t[20];
   int &j = i;
   int *k = &j;

 #pragma omp target update() // expected-warning {{extra tokens at the end of '#pragma omp target update' are ignored}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(alloc) // expected-error {{use of undeclared identifier 'alloc'}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(x)
 #pragma omp target update to(t[:i])
 #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) // le50-error {{expected addressable lvalue in 'to' clause}} 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(argc > 0 ? x : 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(x ?: 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(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)
 #pragma omp target update to(h) // expected-error {{threadprivate variables are not allowed in 'to' clause}} expected-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
 #pragma omp target update to(k), from(k) // expected-error {{variable can appear only once in OpenMP 'target update' construct}} expected-note {{used here}}
 #pragma omp target update to(t), to(t[:5]) // expected-error {{variable can appear only once in OpenMP 'target update' construct}} expected-note {{used here}}
 #pragma omp target update to(da)
 #pragma omp target update to(da[:4])

 #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))) // 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+(m))) // expected-error {{invalid operands to binary expression ('int *' and 'int *')}}
 #pragma omp target update to(*(1+y+y)) // expected-error {{indirection requires pointer operand ('int' invalid)}}
   {}
+
+#pragma omp target update to(id)        // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(id(x))     // 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(id(id(x))) // 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(*id2(&x))  // 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(id2(&x))   // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+
+#pragma omp target update to(reinterpret_cast <int *>(&s10))  // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*reinterpret_cast <int *>(&s10)) // 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(static_cast <S9>(s10))           // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*static_cast <S9 *>(&s10))       // 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(*dynamic_cast <S9 *>(&s10))      // 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(const_cast <S10 *>(&s10))        // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*const_cast <S10 *>(&s10))       // 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((S9)s10)                         // 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'}} le50-error {{expected addressable lvalue in 'to' clause}} le50-error {{expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'}}
+#pragma omp target update to(*((S9 *)&s10))                   // 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'}}
+
   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}}
 }