Hi ABataev, rsmith, hfinkel, fraggamuffin,
Parsing and sema support for #pragma omp target data
http://reviews.llvm.org/D10765
Files:
include/clang-c/Index.h
include/clang/AST/DataRecursiveASTVisitor.h
include/clang/AST/OpenMPClause.h
include/clang/AST/RecursiveASTVisitor.h
include/clang/AST/StmtOpenMP.h
include/clang/Basic/OpenMPKinds.def
include/clang/Basic/StmtNodes.td
include/clang/Sema/Sema.h
include/clang/Serialization/ASTBitCodes.h
lib/AST/Stmt.cpp
lib/AST/StmtPrinter.cpp
lib/AST/StmtProfile.cpp
lib/Basic/OpenMPKinds.cpp
lib/CodeGen/CGStmt.cpp
lib/CodeGen/CGStmtOpenMP.cpp
lib/CodeGen/CodeGenFunction.h
lib/Parse/ParseOpenMP.cpp
lib/Sema/SemaOpenMP.cpp
lib/Sema/TreeTransform.h
lib/Serialization/ASTReaderStmt.cpp
lib/Serialization/ASTWriterStmt.cpp
lib/StaticAnalyzer/Core/ExprEngine.cpp
test/OpenMP/target_data_ast_print.cpp
test/OpenMP/target_data_if_messages.cpp
test/OpenMP/target_data_messages.c
tools/libclang/CIndex.cpp
tools/libclang/CXCursor.cpp
EMAIL PREFERENCES
http://reviews.llvm.org/settings/panel/emailpreferences/
Index: include/clang-c/Index.h
===================================================================
--- include/clang-c/Index.h
+++ include/clang-c/Index.h
@@ -2227,10 +2227,13 @@
/** \brief OpenMP taskwait directive.
*/
- CXCursor_OMPTaskgroupDirective = 254,
+ CXCursor_OMPTaskgroupDirective = 254,
+ /** \brief OpenMP target data directive.
+ */
+ CXCursor_OMPTargetDataDirective = 255,
- CXCursor_LastStmt = CXCursor_OMPTaskgroupDirective,
+ CXCursor_LastStmt = CXCursor_OMPTargetDataDirective,
/**
* \brief Cursor that represents the translation unit itself.
Index: include/clang/AST/DataRecursiveASTVisitor.h
===================================================================
--- include/clang/AST/DataRecursiveASTVisitor.h
+++ include/clang/AST/DataRecursiveASTVisitor.h
@@ -2370,6 +2370,9 @@
DEF_TRAVERSE_STMT(OMPTargetDirective,
{ TRY_TO(TraverseOMPExecutableDirective(S)); })
+DEF_TRAVERSE_STMT(OMPTargetDataDirective,
+ { TRY_TO(TraverseOMPExecutableDirective(S)); })
+
DEF_TRAVERSE_STMT(OMPTeamsDirective,
{ TRY_TO(TraverseOMPExecutableDirective(S)); })
Index: include/clang/AST/OpenMPClause.h
===================================================================
--- include/clang/AST/OpenMPClause.h
+++ include/clang/AST/OpenMPClause.h
@@ -18,6 +18,7 @@
#include "clang/AST/Expr.h"
#include "clang/AST/Stmt.h"
+#include "clang/AST/PrettyPrinter.h"
#include "clang/Basic/OpenMPKinds.h"
#include "clang/Basic/SourceLocation.h"
@@ -127,6 +128,7 @@
varlist_iterator varlist_end() { return getVarRefs().end(); }
varlist_const_iterator varlist_begin() const { return getVarRefs().begin(); }
varlist_const_iterator varlist_end() const { return getVarRefs().end(); }
+ unsigned numberOfVariables() const { return NumVars; }
/// \brief Sets the location of '('.
void setLParenLoc(SourceLocation Loc) { LParenLoc = Loc; }
Index: include/clang/AST/RecursiveASTVisitor.h
===================================================================
--- include/clang/AST/RecursiveASTVisitor.h
+++ include/clang/AST/RecursiveASTVisitor.h
@@ -2403,6 +2403,9 @@
DEF_TRAVERSE_STMT(OMPTargetDirective,
{ TRY_TO(TraverseOMPExecutableDirective(S)); })
+DEF_TRAVERSE_STMT(OMPTargetDataDirective,
+ { TRY_TO(TraverseOMPExecutableDirective(S)); })
+
DEF_TRAVERSE_STMT(OMPTeamsDirective,
{ TRY_TO(TraverseOMPExecutableDirective(S)); })
Index: include/clang/AST/StmtOpenMP.h
===================================================================
--- include/clang/AST/StmtOpenMP.h
+++ include/clang/AST/StmtOpenMP.h
@@ -1798,6 +1798,63 @@
}
};
+/// \brief This represents '#pragma omp target data' directive.
+///
+/// \code
+/// #pragma omp target data device(0) if(a) map(b[:])
+/// \endcode
+/// In this example directive '#pragma omp target data' has clauses 'device'
+/// with the value '0', 'if' with condition 'a' and 'map' with array
+/// section 'b[:]'.
+///
+class OMPTargetDataDirective : public OMPExecutableDirective {
+ friend class ASTStmtReader;
+ /// \brief Build directive with the given start and end location.
+ ///
+ /// \param StartLoc Starting location of the directive kind.
+ /// \param EndLoc Ending Location of the directive.
+ /// \param N The number of clauses.
+ ///
+ OMPTargetDataDirective(SourceLocation StartLoc, SourceLocation EndLoc,
+ unsigned NumClauses)
+ : OMPExecutableDirective(this, OMPTargetDataDirectiveClass,
+ OMPD_target_data, StartLoc, EndLoc, NumClauses, 1) {}
+
+ /// \brief Build an empty directive.
+ ///
+ /// \param N Number of clauses.
+ ///
+ explicit OMPTargetDataDirective(unsigned NumClauses)
+ : OMPExecutableDirective(this, OMPTargetDataDirectiveClass,
+ OMPD_target_data, SourceLocation(),
+ SourceLocation(), NumClauses, 1) {}
+
+public:
+ /// \brief Creates directive with a list of \a Clauses.
+ ///
+ /// \param C AST context.
+ /// \param StartLoc Starting location of the directive kind.
+ /// \param EndLoc Ending Location of the directive.
+ /// \param Clauses List of clauses.
+ /// \param AssociatedStmt Statement, associated with the directive.
+ ///
+ static OMPTargetDataDirective *
+ Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+ ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt);
+
+ /// \brief Creates an empty directive with the place for \a N clauses.
+ ///
+ /// \param C AST context.
+ /// \param N The number of clauses.
+ ///
+ static OMPTargetDataDirective *CreateEmpty(const ASTContext &C, unsigned N,
+ EmptyShell);
+
+ static bool classof(const Stmt *T) {
+ return T->getStmtClass() == OMPTargetDataDirectiveClass;
+ }
+};
+
/// \brief This represents '#pragma omp teams' directive.
///
/// \code
Index: include/clang/Basic/OpenMPKinds.def
===================================================================
--- include/clang/Basic/OpenMPKinds.def
+++ include/clang/Basic/OpenMPKinds.def
@@ -57,6 +57,9 @@
#ifndef OPENMP_TARGET_CLAUSE
# define OPENMP_TARGET_CLAUSE(Name)
#endif
+#ifndef OPENMP_TARGET_DATA_CLAUSE
+# define OPENMP_TARGET_DATA_CLAUSE(Name)
+#endif
#ifndef OPENMP_TEAMS_CLAUSE
# define OPENMP_TEAMS_CLAUSE(Name)
#endif
@@ -93,6 +96,7 @@
OPENMP_DIRECTIVE(atomic)
OPENMP_DIRECTIVE(target)
OPENMP_DIRECTIVE(teams)
+OPENMP_DIRECTIVE_EXT(target_data, "target data")
OPENMP_DIRECTIVE_EXT(parallel_for, "parallel for")
OPENMP_DIRECTIVE_EXT(parallel_for_simd, "parallel for simd")
OPENMP_DIRECTIVE_EXT(parallel_sections, "parallel sections")
@@ -270,6 +274,10 @@
// TODO More clauses for 'target' directive.
OPENMP_TARGET_CLAUSE(if)
+// Clauses allowed for OpenMP directive 'target data'.
+// TODO More clauses for 'target data' directive.
+OPENMP_TARGET_DATA_CLAUSE(if)
+
// Clauses allowed for OpenMP directive 'teams'.
// TODO More clauses for 'teams' directive.
OPENMP_TEAMS_CLAUSE(default)
@@ -294,6 +302,7 @@
#undef OPENMP_TASK_CLAUSE
#undef OPENMP_ATOMIC_CLAUSE
#undef OPENMP_TARGET_CLAUSE
+#undef OPENMP_TARGET_DATA_CLAUSE
#undef OPENMP_TEAMS_CLAUSE
#undef OPENMP_SIMD_CLAUSE
#undef OPENMP_FOR_CLAUSE
Index: include/clang/Basic/StmtNodes.td
===================================================================
--- include/clang/Basic/StmtNodes.td
+++ include/clang/Basic/StmtNodes.td
@@ -204,4 +204,5 @@
def OMPOrderedDirective : DStmt<OMPExecutableDirective>;
def OMPAtomicDirective : DStmt<OMPExecutableDirective>;
def OMPTargetDirective : DStmt<OMPExecutableDirective>;
+def OMPTargetDataDirective : DStmt<OMPExecutableDirective>;
def OMPTeamsDirective : DStmt<OMPExecutableDirective>;
Index: include/clang/Sema/Sema.h
===================================================================
--- include/clang/Sema/Sema.h
+++ include/clang/Sema/Sema.h
@@ -7760,6 +7760,11 @@
StmtResult ActOnOpenMPTargetDirective(ArrayRef<OMPClause *> Clauses,
Stmt *AStmt, SourceLocation StartLoc,
SourceLocation EndLoc);
+ /// \brief Called on well-formed '\#pragma omp target data' after parsing of
+ /// the associated statement.
+ StmtResult ActOnOpenMPTargetDataDirective(ArrayRef<OMPClause *> Clauses,
+ Stmt *AStmt, SourceLocation StartLoc,
+ SourceLocation EndLoc);
/// \brief Called on well-formed '\#pragma omp teams' after parsing of the
/// associated statement.
StmtResult ActOnOpenMPTeamsDirective(ArrayRef<OMPClause *> Clauses,
Index: include/clang/Serialization/ASTBitCodes.h
===================================================================
--- include/clang/Serialization/ASTBitCodes.h
+++ include/clang/Serialization/ASTBitCodes.h
@@ -1395,6 +1395,7 @@
STMT_OMP_ORDERED_DIRECTIVE,
STMT_OMP_ATOMIC_DIRECTIVE,
STMT_OMP_TARGET_DIRECTIVE,
+ STMT_OMP_TARGET_DATA_DIRECTIVE,
STMT_OMP_TEAMS_DIRECTIVE,
STMT_OMP_TASKGROUP_DIRECTIVE,
Index: lib/AST/Stmt.cpp
===================================================================
--- lib/AST/Stmt.cpp
+++ lib/AST/Stmt.cpp
@@ -2185,6 +2185,30 @@
return new (Mem) OMPTargetDirective(NumClauses);
}
+OMPTargetDataDirective *OMPTargetDataDirective::Create(
+ const ASTContext &C, SourceLocation StartLoc, SourceLocation EndLoc,
+ ArrayRef<OMPClause *> Clauses, Stmt *AssociatedStmt) {
+ void *Mem =
+ C.Allocate(llvm::RoundUpToAlignment(sizeof(OMPTargetDataDirective),
+ llvm::alignOf<OMPClause *>()) +
+ sizeof(OMPClause *) * Clauses.size() + sizeof(Stmt *));
+ OMPTargetDataDirective *Dir =
+ new (Mem) OMPTargetDataDirective(StartLoc, EndLoc, Clauses.size());
+ Dir->setClauses(Clauses);
+ Dir->setAssociatedStmt(AssociatedStmt);
+ return Dir;
+}
+
+OMPTargetDataDirective *OMPTargetDataDirective::CreateEmpty(const ASTContext &C,
+ unsigned N,
+ EmptyShell) {
+ void *Mem =
+ C.Allocate(llvm::RoundUpToAlignment(sizeof(OMPTargetDataDirective),
+ llvm::alignOf<OMPClause *>()) +
+ sizeof(OMPClause *) * N + sizeof(Stmt *));
+ return new (Mem) OMPTargetDataDirective(N);
+}
+
OMPTeamsDirective *OMPTeamsDirective::Create(const ASTContext &C,
SourceLocation StartLoc,
SourceLocation EndLoc,
Index: lib/AST/StmtPrinter.cpp
===================================================================
--- lib/AST/StmtPrinter.cpp
+++ lib/AST/StmtPrinter.cpp
@@ -946,6 +946,11 @@
PrintOMPExecutableDirective(Node);
}
+void StmtPrinter::VisitOMPTargetDataDirective(OMPTargetDataDirective *Node) {
+ Indent() << "#pragma omp target data ";
+ PrintOMPExecutableDirective(Node);
+}
+
void StmtPrinter::VisitOMPTeamsDirective(OMPTeamsDirective *Node) {
Indent() << "#pragma omp teams ";
PrintOMPExecutableDirective(Node);
Index: lib/AST/StmtProfile.cpp
===================================================================
--- lib/AST/StmtProfile.cpp
+++ lib/AST/StmtProfile.cpp
@@ -533,6 +533,10 @@
VisitOMPExecutableDirective(S);
}
+void StmtProfiler::VisitOMPTargetDataDirective(const OMPTargetDataDirective *S) {
+ VisitOMPExecutableDirective(S);
+}
+
void StmtProfiler::VisitOMPTeamsDirective(const OMPTeamsDirective *S) {
VisitOMPExecutableDirective(S);
}
Index: lib/Basic/OpenMPKinds.cpp
===================================================================
--- lib/Basic/OpenMPKinds.cpp
+++ lib/Basic/OpenMPKinds.cpp
@@ -328,6 +328,16 @@
break;
}
break;
+ case OMPD_target_data:
+ switch (CKind) {
+#define OPENMP_TARGET_DATA_CLAUSE(Name) \
+ case OMPC_##Name: \
+ return true;
+#include "clang/Basic/OpenMPKinds.def"
+ default:
+ break;
+ }
+ break;
case OMPD_teams:
switch (CKind) {
#define OPENMP_TEAMS_CLAUSE(Name) \
Index: lib/CodeGen/CGStmt.cpp
===================================================================
--- lib/CodeGen/CGStmt.cpp
+++ lib/CodeGen/CGStmt.cpp
@@ -237,6 +237,9 @@
case Stmt::OMPTargetDirectiveClass:
EmitOMPTargetDirective(cast<OMPTargetDirective>(*S));
break;
+ case Stmt::OMPTargetDataDirectiveClass:
+ EmitOMPTargetDataDirective(cast<OMPTargetDataDirective>(*S));
+ break;
case Stmt::OMPTeamsDirectiveClass:
EmitOMPTeamsDirective(cast<OMPTeamsDirective>(*S));
break;
Index: lib/CodeGen/CGStmtOpenMP.cpp
===================================================================
--- lib/CodeGen/CGStmtOpenMP.cpp
+++ lib/CodeGen/CGStmtOpenMP.cpp
@@ -2094,6 +2094,12 @@
llvm_unreachable("CodeGen for 'omp target' is not supported yet.");
}
+// Generate the instructions for '#pragma omp target data' directive.
+void CodeGenFunction::EmitOMPTargetDataDirective(
+ const OMPTargetDataDirective &S) {
+ llvm_unreachable("CodeGen for 'omp target data' is not supported yet.");
+}
+
void CodeGenFunction::EmitOMPTeamsDirective(const OMPTeamsDirective &) {
llvm_unreachable("CodeGen for 'omp teams' is not supported yet.");
}
Index: lib/CodeGen/CodeGenFunction.h
===================================================================
--- lib/CodeGen/CodeGenFunction.h
+++ lib/CodeGen/CodeGenFunction.h
@@ -2198,6 +2198,7 @@
void EmitOMPOrderedDirective(const OMPOrderedDirective &S);
void EmitOMPAtomicDirective(const OMPAtomicDirective &S);
void EmitOMPTargetDirective(const OMPTargetDirective &S);
+ void EmitOMPTargetDataDirective(const OMPTargetDataDirective &S);
void EmitOMPTeamsDirective(const OMPTeamsDirective &S);
/// \brief Emit inner loop of the worksharing/simd construct.
Index: lib/Parse/ParseOpenMP.cpp
===================================================================
--- lib/Parse/ParseOpenMP.cpp
+++ lib/Parse/ParseOpenMP.cpp
@@ -40,6 +40,19 @@
Tok.isAnnotation()
? OMPD_unknown
: getOpenMPDirectiveKind(P.getPreprocessor().getSpelling(Tok));
+
+ switch (DKind) {
+ case OMPD_target: {
+ auto SavedToken = P.getPreprocessor().LookAhead(0);
+ if (!SavedToken.isAnnotation()) {
+ if (P.getPreprocessor().getSpelling(SavedToken) == "data") {
+ DKind = OMPD_target_data;
+ P.ConsumeToken();
+ }
+ }
+ break;
+ }
+ default:
for (unsigned i = 0; i < llvm::array_lengthof(F); ++i) {
if (DKind == F[i][0]) {
Tok = P.getPreprocessor().LookAhead(0);
@@ -53,6 +66,7 @@
}
}
}
+ }
return DKind;
}
@@ -129,8 +143,8 @@
/// 'section' | 'single' | 'master' | 'critical' [ '(' <name> ')' ] |
/// 'parallel for' | 'parallel sections' | 'task' | 'taskyield' |
/// 'barrier' | 'taskwait' | 'flush' | 'ordered' | 'atomic' |
-/// 'for simd' | 'parallel for simd' | 'target' | 'teams' | 'taskgroup'
-/// {clause}
+/// 'for simd' | 'parallel for simd' | 'target' | 'teams' |
+/// 'taskgroup' | 'target data' {clause}
/// annot_pragma_openmp_end
///
StmtResult
@@ -200,6 +214,7 @@
case OMPD_ordered:
case OMPD_atomic:
case OMPD_target:
+ case OMPD_target_data:
case OMPD_teams:
case OMPD_taskgroup: {
ConsumeToken();
@@ -360,7 +375,7 @@
/// lastprivate-clause | reduction-clause | proc_bind-clause |
/// schedule-clause | copyin-clause | copyprivate-clause | untied-clause |
/// mergeable-clause | flush-clause | read-clause | write-clause |
-/// update-clause | capture-clause | seq_cst-clause
+/// update-clause | capture-clause | seq_cst-clause
///
OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind,
OpenMPClauseKind CKind, bool FirstClause) {
Index: lib/Sema/SemaOpenMP.cpp
===================================================================
--- lib/Sema/SemaOpenMP.cpp
+++ lib/Sema/SemaOpenMP.cpp
@@ -1280,6 +1280,7 @@
Params);
break;
}
+ case OMPD_target_data:
case OMPD_target: {
Sema::CapturedParamNameType Params[] = {
std::make_pair(StringRef(), QualType()) // __context with shared vars
@@ -2018,6 +2019,10 @@
Res = ActOnOpenMPTargetDirective(ClausesWithImplicit, AStmt, StartLoc,
EndLoc);
break;
+ case OMPD_target_data:
+ Res = ActOnOpenMPTargetDataDirective(ClausesWithImplicit, AStmt, StartLoc,
+ EndLoc);
+ break;
case OMPD_threadprivate:
llvm_unreachable("OpenMP Directive is not allowed");
case OMPD_unknown:
@@ -4187,6 +4192,16 @@
return OMPTargetDirective::Create(Context, StartLoc, EndLoc, Clauses, AStmt);
}
+StmtResult Sema::ActOnOpenMPTargetDataDirective(ArrayRef<OMPClause *> Clauses,
+ Stmt *AStmt,
+ SourceLocation StartLoc,
+ SourceLocation EndLoc) {
+ getCurFunction()->setHasBranchProtectedScope();
+
+ return OMPTargetDataDirective::Create(Context, StartLoc, EndLoc, Clauses,
+ AStmt);
+}
+
StmtResult Sema::ActOnOpenMPTeamsDirective(ArrayRef<OMPClause *> Clauses,
Stmt *AStmt, SourceLocation StartLoc,
SourceLocation EndLoc) {
@@ -6417,4 +6432,3 @@
return OMPDependClause::Create(Context, StartLoc, LParenLoc, EndLoc, DepKind,
DepLoc, ColonLoc, Vars);
}
-
Index: lib/Sema/TreeTransform.h
===================================================================
--- lib/Sema/TreeTransform.h
+++ lib/Sema/TreeTransform.h
@@ -6952,6 +6952,17 @@
}
template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOMPTargetDataDirective(
+ OMPTargetDataDirective *D) {
+ DeclarationNameInfo DirName;
+ getDerived().getSema().StartOpenMPDSABlock(OMPD_target_data, DirName, nullptr,
+ D->getLocStart());
+ StmtResult Res = getDerived().TransformOMPExecutableDirective(D);
+ getDerived().getSema().EndOpenMPDSABlock(Res.get());
+ return Res;
+}
+
+template <typename Derived>
StmtResult
TreeTransform<Derived>::TransformOMPTeamsDirective(OMPTeamsDirective *D) {
DeclarationNameInfo DirName;
Index: lib/Serialization/ASTReaderStmt.cpp
===================================================================
--- lib/Serialization/ASTReaderStmt.cpp
+++ lib/Serialization/ASTReaderStmt.cpp
@@ -2242,6 +2242,12 @@
VisitOMPExecutableDirective(D);
}
+void ASTStmtReader::VisitOMPTargetDataDirective(OMPTargetDataDirective *D) {
+ VisitStmt(D);
+ ++Idx;
+ VisitOMPExecutableDirective(D);
+}
+
void ASTStmtReader::VisitOMPTeamsDirective(OMPTeamsDirective *D) {
VisitStmt(D);
// The NumClauses field was read in ReadStmtFromStream.
@@ -2847,6 +2853,11 @@
Context, Record[ASTStmtReader::NumStmtFields], Empty);
break;
+ case STMT_OMP_TARGET_DATA_DIRECTIVE:
+ S = OMPTargetDataDirective::CreateEmpty(
+ Context, Record[ASTStmtReader::NumStmtFields], Empty);
+ break;
+
case STMT_OMP_TEAMS_DIRECTIVE:
S = OMPTeamsDirective::CreateEmpty(
Context, Record[ASTStmtReader::NumStmtFields], Empty);
Index: lib/Serialization/ASTWriterStmt.cpp
===================================================================
--- lib/Serialization/ASTWriterStmt.cpp
+++ lib/Serialization/ASTWriterStmt.cpp
@@ -2063,6 +2063,13 @@
Code = serialization::STMT_OMP_TARGET_DIRECTIVE;
}
+void ASTStmtWriter::VisitOMPTargetDataDirective(OMPTargetDataDirective *D) {
+ VisitStmt(D);
+ Record.push_back(D->getNumClauses());
+ VisitOMPExecutableDirective(D);
+ Code = serialization::STMT_OMP_TARGET_DATA_DIRECTIVE;
+}
+
void ASTStmtWriter::VisitOMPTaskyieldDirective(OMPTaskyieldDirective *D) {
VisitStmt(D);
VisitOMPExecutableDirective(D);
Index: lib/StaticAnalyzer/Core/ExprEngine.cpp
===================================================================
--- lib/StaticAnalyzer/Core/ExprEngine.cpp
+++ lib/StaticAnalyzer/Core/ExprEngine.cpp
@@ -821,6 +821,7 @@
case Stmt::OMPOrderedDirectiveClass:
case Stmt::OMPAtomicDirectiveClass:
case Stmt::OMPTargetDirectiveClass:
+ case Stmt::OMPTargetDataDirectiveClass:
case Stmt::OMPTeamsDirectiveClass:
llvm_unreachable("Stmt should not be in analyzer evaluation loop");
Index: test/OpenMP/target_data_ast_print.cpp
===================================================================
--- /dev/null
+++ test/OpenMP/target_data_ast_print.cpp
@@ -0,0 +1,26 @@
+// RUN: %clang_cc1 -verify -fopenmp -ast-print %s | FileCheck %s
+// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -emit-pch -o %t %s
+// RUN: %clang_cc1 -fopenmp -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print | FileCheck %s
+// expected-no-diagnostics
+
+#ifndef HEADER
+#define HEADER
+
+void foo() {}
+
+int main (int argc, char **argv) {
+ int b = argc, c, d, e, f, g;
+ static int a;
+// CHECK: static int a;
+#pragma omp target data
+// CHECK: #pragma omp target data
+ a=2;
+// CHECK-NEXT: a = 2;
+#pragma omp target data if(b)
+// CHECK: #pragma omp target data if(b)
+ foo();
+// CHECK-NEXT: foo();
+ return (0);
+}
+
+#endif
Index: test/OpenMP/target_data_if_messages.cpp
===================================================================
--- /dev/null
+++ test/OpenMP/target_data_if_messages.cpp
@@ -0,0 +1,26 @@
+// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s
+
+void foo() {
+}
+
+bool foobool(int argc) {
+ return argc;
+}
+
+struct S1; // expected-note {{declared here}}
+
+int main(int argc, char **argv) {
+ #pragma omp target data if // expected-error {{expected '(' after 'if'}}
+ #pragma omp target data if ( // expected-error {{expected expression}} expected-error {{expected ')'}} expected-note {{to match this '('}}
+ #pragma omp target data if () // expected-error {{expected expression}}
+ #pragma omp target data if (argc // expected-error {{expected ')'}} expected-note {{to match this '('}}
+ #pragma omp target data if (argc)) // expected-warning {{extra tokens at the end of '#pragma omp target data' are ignored}}
+ #pragma omp target data if (argc > 0 ? argv[1] : argv[2])
+ #pragma omp target data if (argc + argc)
+ #pragma omp target data if (foobool(argc)), if (true) // expected-error {{directive '#pragma omp target data' cannot contain more than one 'if' clause}}
+ #pragma omp target data if (S1) // expected-error {{'S1' does not refer to a value}}
+ #pragma omp target data if (argv[1]=2) // expected-error {{expected ')'}} expected-note {{to match this '('}}
+ foo();
+
+ return 0;
+}
Index: test/OpenMP/target_data_messages.c
===================================================================
--- /dev/null
+++ test/OpenMP/target_data_messages.c
@@ -0,0 +1,19 @@
+// RUN: %clang_cc1 -triple x86_64-apple-macos10.7.0 -verify -fopenmp -ferror-limit 100 -o - %s
+
+void foo() { }
+
+int main(int argc, char **argv) {
+ L1:
+ foo();
+ #pragma omp target data
+ {
+ foo();
+ goto L1; // expected-error {{use of undeclared label 'L1'}}
+ }
+ goto L2; // expected-error {{use of undeclared label 'L2'}}
+ #pragma omp target data
+ L2:
+ foo();
+
+ return 0;
+}
Index: tools/libclang/CIndex.cpp
===================================================================
--- tools/libclang/CIndex.cpp
+++ tools/libclang/CIndex.cpp
@@ -1884,6 +1884,7 @@
void VisitOMPOrderedDirective(const OMPOrderedDirective *D);
void VisitOMPAtomicDirective(const OMPAtomicDirective *D);
void VisitOMPTargetDirective(const OMPTargetDirective *D);
+ void VisitOMPTargetDataDirective(const OMPTargetDataDirective *D);
void VisitOMPTeamsDirective(const OMPTeamsDirective *D);
private:
@@ -2503,6 +2504,11 @@
VisitOMPExecutableDirective(D);
}
+void EnqueueVisitor::VisitOMPTargetDataDirective(const
+ OMPTargetDataDirective *D) {
+ VisitOMPExecutableDirective(D);
+}
+
void EnqueueVisitor::VisitOMPTeamsDirective(const OMPTeamsDirective *D) {
VisitOMPExecutableDirective(D);
}
@@ -4281,6 +4287,8 @@
return cxstring::createRef("OMPAtomicDirective");
case CXCursor_OMPTargetDirective:
return cxstring::createRef("OMPTargetDirective");
+ case CXCursor_OMPTargetDataDirective:
+ return cxstring::createRef("OMPTargetDataDirective");
case CXCursor_OMPTeamsDirective:
return cxstring::createRef("OMPTeamsDirective");
case CXCursor_OverloadCandidate:
Index: tools/libclang/CXCursor.cpp
===================================================================
--- tools/libclang/CXCursor.cpp
+++ tools/libclang/CXCursor.cpp
@@ -585,6 +585,9 @@
case Stmt::OMPTargetDirectiveClass:
K = CXCursor_OMPTargetDirective;
break;
+ case Stmt::OMPTargetDataDirectiveClass:
+ K = CXCursor_OMPTargetDataDirective;
+ break;
case Stmt::OMPTeamsDirectiveClass:
K = CXCursor_OMPTeamsDirective;
break;
_______________________________________________
cfe-commits mailing list
cfe-commits@cs.uiuc.edu
http://lists.cs.uiuc.edu/mailman/listinfo/cfe-commits
