r275271 - [OpenMP] Initial implementation of parse+sema for clause use_device_ptr of 'target data'
Carlo Bertolli via cfe-commits
cfe-commits at lists.llvm.org
Wed Jul 13 08:37:17 PDT 2016
Author: cbertol
Date: Wed Jul 13 10:37:16 2016
New Revision: 275271
URL: http://llvm.org/viewvc/llvm-project?rev=275271&view=rev
Log:
[OpenMP] Initial implementation of parse+sema for clause use_device_ptr of 'target data'
http://reviews.llvm.org/D21904
This patch is similar to the implementation of 'private' clause: it adds a list of private pointers to be used within the target data region to store the device pointers returned by the runtime.
Please refer to the following document for a full description of what the runtime witll return in this case (page 10 and 11):
https://github.com/clang-omp/OffloadingDesign
I am happy to answer any question related to the runtime interface to help reviewing this patch.
Added:
cfe/trunk/test/OpenMP/target_data_use_device_ptr_ast_print.cpp
cfe/trunk/test/OpenMP/target_data_use_device_ptr_messages.cpp
Modified:
cfe/trunk/include/clang/AST/OpenMPClause.h
cfe/trunk/include/clang/AST/RecursiveASTVisitor.h
cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td
cfe/trunk/include/clang/Basic/OpenMPKinds.def
cfe/trunk/include/clang/Sema/Sema.h
cfe/trunk/lib/AST/OpenMPClause.cpp
cfe/trunk/lib/AST/StmtPrinter.cpp
cfe/trunk/lib/AST/StmtProfile.cpp
cfe/trunk/lib/Basic/OpenMPKinds.cpp
cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp
cfe/trunk/lib/Parse/ParseOpenMP.cpp
cfe/trunk/lib/Sema/SemaOpenMP.cpp
cfe/trunk/lib/Sema/TreeTransform.h
cfe/trunk/lib/Serialization/ASTReaderStmt.cpp
cfe/trunk/lib/Serialization/ASTWriterStmt.cpp
cfe/trunk/tools/libclang/CIndex.cpp
Modified: cfe/trunk/include/clang/AST/OpenMPClause.h
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/AST/OpenMPClause.h?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/include/clang/AST/OpenMPClause.h (original)
+++ cfe/trunk/include/clang/AST/OpenMPClause.h Wed Jul 13 10:37:16 2016
@@ -4217,6 +4217,71 @@ public:
reinterpret_cast<Stmt **>(varlist_end()));
}
};
+
+/// This represents clause 'use_device_ptr' in the '#pragma omp ...'
+/// directives.
+///
+/// \code
+/// #pragma omp target data use_device_ptr(a,b)
+/// \endcode
+/// In this example directive '#pragma omp target data' has clause
+/// 'use_device_ptr' with the variables 'a' and 'b'.
+///
+class OMPUseDevicePtrClause final
+ : public OMPVarListClause<OMPUseDevicePtrClause>,
+ private llvm::TrailingObjects<OMPUseDevicePtrClause, Expr *> {
+ friend TrailingObjects;
+ friend OMPVarListClause;
+ friend class OMPClauseReader;
+ /// Build clause with number of variables \a N.
+ ///
+ /// \param StartLoc Starting location of the clause.
+ /// \param LParenLoc Location of '('.
+ /// \param EndLoc Ending location of the clause.
+ /// \param N Number of the variables in the clause.
+ ///
+ OMPUseDevicePtrClause(SourceLocation StartLoc, SourceLocation LParenLoc,
+ SourceLocation EndLoc, unsigned N)
+ : OMPVarListClause<OMPUseDevicePtrClause>(OMPC_use_device_ptr, StartLoc,
+ LParenLoc, EndLoc, N) {}
+
+ /// \brief Build an empty clause.
+ ///
+ /// \param N Number of variables.
+ ///
+ explicit OMPUseDevicePtrClause(unsigned N)
+ : OMPVarListClause<OMPUseDevicePtrClause>(
+ OMPC_use_device_ptr, SourceLocation(), SourceLocation(),
+ SourceLocation(), N) {}
+
+public:
+ /// Creates clause with a list of variables \a VL.
+ ///
+ /// \param C AST context.
+ /// \param StartLoc Starting location of the clause.
+ /// \param LParenLoc Location of '('.
+ /// \param EndLoc Ending location of the clause.
+ /// \param VL List of references to the variables.
+ ///
+ static OMPUseDevicePtrClause *
+ Create(const ASTContext &C, SourceLocation StartLoc, SourceLocation LParenLoc,
+ SourceLocation EndLoc, ArrayRef<Expr *> VL);
+ /// Creates an empty clause with the place for \a N variables.
+ ///
+ /// \param C AST context.
+ /// \param N The number of variables.
+ ///
+ static OMPUseDevicePtrClause *CreateEmpty(const ASTContext &C, unsigned N);
+
+ child_range children() {
+ return child_range(reinterpret_cast<Stmt **>(varlist_begin()),
+ reinterpret_cast<Stmt **>(varlist_end()));
+ }
+
+ static bool classof(const OMPClause *T) {
+ return T->getClauseKind() == OMPC_use_device_ptr;
+ }
+};
} // end namespace clang
#endif // LLVM_CLANG_AST_OPENMPCLAUSE_H
Modified: cfe/trunk/include/clang/AST/RecursiveASTVisitor.h
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/AST/RecursiveASTVisitor.h?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/include/clang/AST/RecursiveASTVisitor.h (original)
+++ cfe/trunk/include/clang/AST/RecursiveASTVisitor.h Wed Jul 13 10:37:16 2016
@@ -2952,6 +2952,13 @@ bool RecursiveASTVisitor<Derived>::Visit
return true;
}
+template <typename Derived>
+bool RecursiveASTVisitor<Derived>::VisitOMPUseDevicePtrClause(
+ OMPUseDevicePtrClause *C) {
+ TRY_TO(VisitOMPClauseList(C));
+ return true;
+}
+
// FIXME: look at the following tricky-seeming exprs to see if we
// need to recurse on anything. These are ones that have methods
// returning decls or qualtypes or nestednamespecifier -- though I'm
Modified: cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td (original)
+++ cfe/trunk/include/clang/Basic/DiagnosticSemaKinds.td Wed Jul 13 10:37:16 2016
@@ -8330,6 +8330,8 @@ def err_omp_expected_int_param : Error<
"expected a reference to an integer-typed parameter">;
def err_omp_at_least_one_motion_clause_required : Error<
"expected at least one 'to' clause or 'from' clause specified to '#pragma omp target update'">;
+def err_omp_usedeviceptr_not_a_pointer : Error<
+ "expected pointer or reference to pointer in 'use_device_ptr' clause">;
def warn_omp_nesting_simd : Warning<
"OpenMP only allows an ordered construct with the simd clause nested in a simd construct">,
InGroup<SourceUsesOpenMP>;
Modified: cfe/trunk/include/clang/Basic/OpenMPKinds.def
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Basic/OpenMPKinds.def?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/include/clang/Basic/OpenMPKinds.def (original)
+++ cfe/trunk/include/clang/Basic/OpenMPKinds.def Wed Jul 13 10:37:16 2016
@@ -224,6 +224,7 @@ OPENMP_CLAUSE(dist_schedule, OMPDistSche
OPENMP_CLAUSE(defaultmap, OMPDefaultmapClause)
OPENMP_CLAUSE(to, OMPToClause)
OPENMP_CLAUSE(from, OMPFromClause)
+OPENMP_CLAUSE(use_device_ptr, OMPUseDevicePtrClause)
// Clauses allowed for OpenMP directive 'parallel'.
OPENMP_PARALLEL_CLAUSE(if)
@@ -408,6 +409,7 @@ OPENMP_TARGET_CLAUSE(firstprivate)
OPENMP_TARGET_DATA_CLAUSE(if)
OPENMP_TARGET_DATA_CLAUSE(device)
OPENMP_TARGET_DATA_CLAUSE(map)
+OPENMP_TARGET_DATA_CLAUSE(use_device_ptr)
// Clauses allowed for OpenMP directive 'target enter data'.
OPENMP_TARGET_ENTER_DATA_CLAUSE(if)
Modified: cfe/trunk/include/clang/Sema/Sema.h
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/include/clang/Sema/Sema.h?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/include/clang/Sema/Sema.h (original)
+++ cfe/trunk/include/clang/Sema/Sema.h Wed Jul 13 10:37:16 2016
@@ -8473,6 +8473,11 @@ public:
SourceLocation StartLoc,
SourceLocation LParenLoc,
SourceLocation EndLoc);
+ /// Called on well-formed 'use_device_ptr' clause.
+ OMPClause *ActOnOpenMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc);
/// \brief The kind of conversion being performed.
enum CheckedConversionKind {
Modified: cfe/trunk/lib/AST/OpenMPClause.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/AST/OpenMPClause.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/AST/OpenMPClause.cpp (original)
+++ cfe/trunk/lib/AST/OpenMPClause.cpp Wed Jul 13 10:37:16 2016
@@ -89,6 +89,7 @@ const OMPClauseWithPreInit *OMPClauseWit
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
break;
}
@@ -152,6 +153,7 @@ const OMPClauseWithPostUpdate *OMPClause
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
break;
}
@@ -727,3 +729,21 @@ OMPFromClause *OMPFromClause::CreateEmpt
return new (Mem) OMPFromClause(NumVars, NumUniqueDeclarations,
NumComponentLists, NumComponents);
}
+
+OMPUseDevicePtrClause *OMPUseDevicePtrClause::Create(const ASTContext &C,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc,
+ ArrayRef<Expr *> VL) {
+ void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(VL.size()));
+ OMPUseDevicePtrClause *Clause =
+ new (Mem) OMPUseDevicePtrClause(StartLoc, LParenLoc, EndLoc, VL.size());
+ Clause->setVarRefs(VL);
+ return Clause;
+}
+
+OMPUseDevicePtrClause *OMPUseDevicePtrClause::CreateEmpty(const ASTContext &C,
+ unsigned N) {
+ void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(N));
+ return new (Mem) OMPUseDevicePtrClause(N);
+}
Modified: cfe/trunk/lib/AST/StmtPrinter.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/AST/StmtPrinter.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/AST/StmtPrinter.cpp (original)
+++ cfe/trunk/lib/AST/StmtPrinter.cpp Wed Jul 13 10:37:16 2016
@@ -947,6 +947,14 @@ void OMPClausePrinter::VisitOMPDefaultma
Node->getDefaultmapKind());
OS << ")";
}
+
+void OMPClausePrinter::VisitOMPUseDevicePtrClause(OMPUseDevicePtrClause *Node) {
+ if (!Node->varlist_empty()) {
+ OS << "use_device_ptr";
+ VisitOMPClauseList(Node, '(');
+ OS << ")";
+ }
+}
}
//===----------------------------------------------------------------------===//
Modified: cfe/trunk/lib/AST/StmtProfile.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/AST/StmtProfile.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/AST/StmtProfile.cpp (original)
+++ cfe/trunk/lib/AST/StmtProfile.cpp Wed Jul 13 10:37:16 2016
@@ -529,6 +529,10 @@ void OMPClauseProfiler::VisitOMPToClause
void OMPClauseProfiler::VisitOMPFromClause(const OMPFromClause *C) {
VisitOMPClauseList(C);
}
+void OMPClauseProfiler::VisitOMPUseDevicePtrClause(
+ const OMPUseDevicePtrClause *C) {
+ VisitOMPClauseList(C);
+}
}
void
Modified: cfe/trunk/lib/Basic/OpenMPKinds.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Basic/OpenMPKinds.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Basic/OpenMPKinds.cpp (original)
+++ cfe/trunk/lib/Basic/OpenMPKinds.cpp Wed Jul 13 10:37:16 2016
@@ -164,6 +164,7 @@ unsigned clang::getOpenMPSimpleClauseTyp
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
break;
}
llvm_unreachable("Invalid OpenMP simple clause kind");
@@ -301,6 +302,7 @@ const char *clang::getOpenMPSimpleClause
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
break;
}
llvm_unreachable("Invalid OpenMP simple clause kind");
Modified: cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp (original)
+++ cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp Wed Jul 13 10:37:16 2016
@@ -3186,6 +3186,7 @@ static void EmitOMPAtomicExpr(CodeGenFun
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
llvm_unreachable("Clause is not allowed in 'omp atomic'.");
}
}
Modified: cfe/trunk/lib/Parse/ParseOpenMP.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Parse/ParseOpenMP.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Parse/ParseOpenMP.cpp (original)
+++ cfe/trunk/lib/Parse/ParseOpenMP.cpp Wed Jul 13 10:37:16 2016
@@ -1183,6 +1183,7 @@ OMPClause *Parser::ParseOpenMPClause(Ope
case OMPC_map:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
Clause = ParseOpenMPVarListClause(DKind, CKind);
break;
case OMPC_unknown:
@@ -1747,6 +1748,8 @@ bool Parser::ParseOpenMPVarList(OpenMPDi
/// 'to' '(' list ')'
/// from-clause:
/// 'from' '(' list ')'
+/// use_device_ptr-clause:
+/// 'use_device_ptr' '(' list ')'
///
/// For 'linear' clause linear-list may have the following forms:
/// list
Modified: cfe/trunk/lib/Sema/SemaOpenMP.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/SemaOpenMP.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Sema/SemaOpenMP.cpp (original)
+++ cfe/trunk/lib/Sema/SemaOpenMP.cpp Wed Jul 13 10:37:16 2016
@@ -7269,6 +7269,7 @@ OMPClause *Sema::ActOnOpenMPSingleExprCl
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
@@ -7555,6 +7556,7 @@ OMPClause *Sema::ActOnOpenMPSimpleClause
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
@@ -7708,6 +7710,7 @@ OMPClause *Sema::ActOnOpenMPSingleExprWi
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
@@ -7895,6 +7898,7 @@ OMPClause *Sema::ActOnOpenMPClause(OpenM
case OMPC_uniform:
case OMPC_to:
case OMPC_from:
+ case OMPC_use_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
@@ -8014,6 +8018,9 @@ OMPClause *Sema::ActOnOpenMPVarListClaus
case OMPC_from:
Res = ActOnOpenMPFromClause(VarList, StartLoc, LParenLoc, EndLoc);
break;
+ case OMPC_use_device_ptr:
+ Res = ActOnOpenMPUseDevicePtrClause(VarList, StartLoc, LParenLoc, EndLoc);
+ break;
case OMPC_if:
case OMPC_final:
case OMPC_num_threads:
@@ -11616,3 +11623,39 @@ OMPClause *Sema::ActOnOpenMPFromClause(A
MVLI.ProcessedVarList, MVLI.VarBaseDeclarations,
MVLI.VarComponents);
}
+
+OMPClause *Sema::ActOnOpenMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc) {
+ SmallVector<Expr *, 8> Vars;
+ for (auto &RefExpr : VarList) {
+ assert(RefExpr && "NULL expr in OpenMP use_device_ptr clause.");
+ SourceLocation ELoc;
+ SourceRange ERange;
+ Expr *SimpleRefExpr = RefExpr;
+ auto Res = getPrivateItem(*this, SimpleRefExpr, ELoc, ERange);
+ if (Res.second) {
+ // It will be analyzed later.
+ Vars.push_back(RefExpr);
+ }
+ ValueDecl *D = Res.first;
+ if (!D)
+ continue;
+
+ QualType Type = D->getType();
+ // item should be a pointer or reference to pointer
+ if (!Type.getNonReferenceType()->isPointerType()) {
+ Diag(ELoc, diag::err_omp_usedeviceptr_not_a_pointer)
+ << 0 << RefExpr->getSourceRange();
+ continue;
+ }
+ Vars.push_back(RefExpr->IgnoreParens());
+ }
+
+ if (Vars.empty())
+ return nullptr;
+
+ return OMPUseDevicePtrClause::Create(Context, StartLoc, LParenLoc, EndLoc,
+ Vars);
+}
Modified: cfe/trunk/lib/Sema/TreeTransform.h
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/TreeTransform.h?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Sema/TreeTransform.h (original)
+++ cfe/trunk/lib/Sema/TreeTransform.h Wed Jul 13 10:37:16 2016
@@ -1781,6 +1781,18 @@ public:
EndLoc);
}
+ /// Build a new OpenMP 'use_device_ptr' clause.
+ ///
+ /// By default, performs semantic analysis to build the new OpenMP clause.
+ /// Subclasses may override this routine to provide different behavior.
+ OMPClause *RebuildOMPUseDevicePtrClause(ArrayRef<Expr *> VarList,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc) {
+ return getSema().ActOnOpenMPUseDevicePtrClause(VarList, StartLoc, LParenLoc,
+ EndLoc);
+ }
+
/// \brief Rebuild the operand to an Objective-C \@synchronized statement.
///
/// By default, performs semantic analysis to build the new statement.
@@ -8090,6 +8102,21 @@ OMPClause *TreeTransform<Derived>::Trans
C->getLParenLoc(), C->getLocEnd());
}
+template <typename Derived>
+OMPClause *TreeTransform<Derived>::TransformOMPUseDevicePtrClause(
+ OMPUseDevicePtrClause *C) {
+ llvm::SmallVector<Expr *, 16> Vars;
+ Vars.reserve(C->varlist_size());
+ for (auto *VE : C->varlists()) {
+ ExprResult EVar = getDerived().TransformExpr(cast<Expr>(VE));
+ if (EVar.isInvalid())
+ return nullptr;
+ Vars.push_back(EVar.get());
+ }
+ return getDerived().RebuildOMPUseDevicePtrClause(
+ Vars, C->getLocStart(), C->getLParenLoc(), C->getLocEnd());
+}
+
//===----------------------------------------------------------------------===//
// Expression transformation
//===----------------------------------------------------------------------===//
Modified: cfe/trunk/lib/Serialization/ASTReaderStmt.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Serialization/ASTReaderStmt.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Serialization/ASTReaderStmt.cpp (original)
+++ cfe/trunk/lib/Serialization/ASTReaderStmt.cpp Wed Jul 13 10:37:16 2016
@@ -1922,6 +1922,9 @@ OMPClause *OMPClauseReader::readClause()
NumComponents);
break;
}
+ case OMPC_use_device_ptr:
+ C = OMPUseDevicePtrClause::CreateEmpty(Context, Record[Idx++]);
+ break;
}
Visit(C);
C->setLocStart(Reader->ReadSourceLocation(Record, Idx));
@@ -2439,6 +2442,17 @@ void OMPClauseReader::VisitOMPFromClause
C->setComponents(Components, ListSizes);
}
+void OMPClauseReader::VisitOMPUseDevicePtrClause(OMPUseDevicePtrClause *C) {
+ C->setLParenLoc(Reader->ReadSourceLocation(Record, Idx));
+ unsigned NumVars = C->varlist_size();
+ SmallVector<Expr *, 16> Vars;
+ Vars.reserve(NumVars);
+ for (unsigned i = 0; i != NumVars; ++i)
+ Vars.push_back(Reader->Reader.ReadSubExpr());
+ C->setVarRefs(Vars);
+ Vars.clear();
+}
+
//===----------------------------------------------------------------------===//
// OpenMP Directives.
//===----------------------------------------------------------------------===//
Modified: cfe/trunk/lib/Serialization/ASTWriterStmt.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Serialization/ASTWriterStmt.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/lib/Serialization/ASTWriterStmt.cpp (original)
+++ cfe/trunk/lib/Serialization/ASTWriterStmt.cpp Wed Jul 13 10:37:16 2016
@@ -2140,6 +2140,14 @@ void OMPClauseWriter::VisitOMPFromClause
}
}
+void OMPClauseWriter::VisitOMPUseDevicePtrClause(OMPUseDevicePtrClause *C) {
+ Record.push_back(C->varlist_size());
+ Record.AddSourceLocation(C->getLParenLoc());
+ for (auto *VE : C->varlists()) {
+ Record.AddStmt(VE);
+ }
+}
+
//===----------------------------------------------------------------------===//
// OpenMP Directives.
//===----------------------------------------------------------------------===//
Added: cfe/trunk/test/OpenMP/target_data_use_device_ptr_ast_print.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_data_use_device_ptr_ast_print.cpp?rev=275271&view=auto
==============================================================================
--- cfe/trunk/test/OpenMP/target_data_use_device_ptr_ast_print.cpp (added)
+++ cfe/trunk/test/OpenMP/target_data_use_device_ptr_ast_print.cpp Wed Jul 13 10:37:16 2016
@@ -0,0 +1,154 @@
+// RxUN: %clang_cc1 -verify -fopenmp -std=c++11 -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
+
+struct ST {
+ int *a;
+};
+struct SA {
+ int i, j;
+ int *k = &j;
+ int *&z = k;
+ void func(int arg) {
+#pragma omp target data map(tofrom: i) use_device_ptr(k)
+ {}
+#pragma omp target data map(tofrom: i) use_device_ptr(z)
+ {}
+ return;
+ }
+};
+// CHECK: struct SA
+// CHECK: void func(
+// CHECK: #pragma omp target data map(tofrom: this->i) use_device_ptr(this->k)
+// CHECK: #pragma omp target data map(tofrom: this->i) use_device_ptr(this->z)
+struct SB {
+ unsigned A;
+ unsigned B;
+ float Arr[100];
+ float *Ptr;
+ float *foo() {
+ return &Arr[0];
+ }
+};
+
+struct SC {
+ unsigned A : 2;
+ unsigned B : 3;
+ unsigned C;
+ unsigned D;
+ float Arr[100];
+ SB S;
+ SB ArrS[100];
+ SB *PtrS;
+ SB *&RPtrS;
+ float *Ptr;
+
+ SC(SB *&_RPtrS) : RPtrS(_RPtrS) {}
+};
+
+union SD {
+ unsigned A;
+ float B;
+};
+
+struct S1;
+extern S1 a;
+class S2 {
+ mutable int a;
+public:
+ S2():a(0) { }
+ S2(S2 &s2):a(s2.a) { }
+ static float S2s;
+ static const float S2sc;
+};
+const float S2::S2sc = 0;
+const S2 b;
+const S2 ba[5];
+class S3 {
+ int a;
+public:
+ S3():a(0) { }
+ S3(S3 &s3):a(s3.a) { }
+};
+const S3 c;
+const S3 ca[5];
+extern const int f;
+class S4 {
+ int a;
+ S4();
+ S4(const S4 &s4);
+public:
+ S4(int v):a(v) { }
+};
+class S5 {
+ int a;
+ S5():a(0) {}
+ S5(const S5 &s5):a(s5.a) { }
+public:
+ S5(int v):a(v) { }
+};
+
+S3 h;
+#pragma omp threadprivate(h)
+
+typedef int from;
+
+template <typename T>
+T tmain(T argc) {
+ T i;
+ T &j = i;
+ T *k = &j;
+ T *&z = k;
+#pragma omp target data map(tofrom: i) use_device_ptr(k)
+ {}
+#pragma omp target data map(tofrom: i) use_device_ptr(z)
+ {}
+ return 0;
+}
+
+// CHECK: template <typename T = int> int tmain(int argc) {
+// CHECK-NEXT: int i;
+// CHECK-NEXT: int &j = i;
+// CHECK-NEXT: int *k = &j;
+// CHECK-NEXT: int *&z = k;
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(k)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(z)
+
+// CHECK: template <typename T = int *> int *tmain(int *argc) {
+// CHECK-NEXT: int *i;
+// CHECK-NEXT: int *&j = i;
+// CHECK-NEXT: int **k = &j;
+// CHECK-NEXT: int **&z = k;
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(k)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(z)
+
+// CHECK-LABEL: int main(int argc, char **argv) {
+int main(int argc, char **argv) {
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+// CHECK-NEXT: int i;
+// CHECK-NEXT: int &j = i;
+// CHECK-NEXT: int *k = &j;
+// CHECK-NEXT: int *&z = k;
+#pragma omp target data map(tofrom: i) use_device_ptr(k)
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(k)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target data map(tofrom: i) use_device_ptr(z)
+// CHECK-NEXT: #pragma omp target data map(tofrom: i) use_device_ptr(z)
+ {}
+ return tmain<int>(argc) + (*tmain<int*>(&argc));
+}
+
+#endif
Added: cfe/trunk/test/OpenMP/target_data_use_device_ptr_messages.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_data_use_device_ptr_messages.cpp?rev=275271&view=auto
==============================================================================
--- cfe/trunk/test/OpenMP/target_data_use_device_ptr_messages.cpp (added)
+++ cfe/trunk/test/OpenMP/target_data_use_device_ptr_messages.cpp Wed Jul 13 10:37:16 2016
@@ -0,0 +1,206 @@
+// RUN: %clang_cc1 -std=c++11 -verify -fopenmp -ferror-limit 200 %s
+struct ST {
+ int *a;
+};
+struct SA {
+ const int d = 5;
+ const int da[5] = { 0 };
+ ST e;
+ ST g[10];
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+ void func(int arg) {
+#pragma omp target data map(i) use_device_ptr // expected-error {{expected '(' after 'use_device_ptr'}}
+ {}
+#pragma omp target data map(i) use_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target data map(i) use_device_ptr(arg // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(i) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(j) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(z) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(aa) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(e) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(g) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k,i,j) // expected-error2 {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(d) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(da) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+ return;
+ }
+};
+struct SB {
+ unsigned A;
+ unsigned B;
+ float Arr[100];
+ float *Ptr;
+ float *foo() {
+ return &Arr[0];
+ }
+};
+
+struct SC {
+ unsigned A : 2;
+ unsigned B : 3;
+ unsigned C;
+ unsigned D;
+ float Arr[100];
+ SB S;
+ SB ArrS[100];
+ SB *PtrS;
+ SB *&RPtrS;
+ float *Ptr;
+
+ SC(SB *&_RPtrS) : RPtrS(_RPtrS) {}
+};
+
+union SD {
+ unsigned A;
+ float B;
+};
+
+struct S1;
+extern S1 a;
+class S2 {
+ mutable int a;
+public:
+ S2():a(0) { }
+ S2(S2 &s2):a(s2.a) { }
+ static float S2s;
+ static const float S2sc;
+};
+const float S2::S2sc = 0;
+const S2 b;
+const S2 ba[5];
+class S3 {
+ int a;
+public:
+ S3():a(0) { }
+ S3(S3 &s3):a(s3.a) { }
+};
+const S3 c;
+const S3 ca[5];
+extern const int f;
+class S4 {
+ int a;
+ S4();
+ S4(const S4 &s4);
+public:
+ S4(int v):a(v) { }
+};
+class S5 {
+ int a;
+ S5():a(0) {}
+ S5(const S5 &s5):a(s5.a) { }
+public:
+ S5(int v):a(v) { }
+};
+
+S3 h;
+#pragma omp threadprivate(h)
+
+typedef int from;
+
+template <typename T, int I>
+T tmain(T argc) {
+ const T d = 5;
+ const T da[5] = { 0 };
+ S4 e(4);
+ S5 g(5);
+ T i;
+ T &j = i;
+ T *k = &j;
+ T *&z = k;
+ T aa[10];
+#pragma omp target data map(i) use_device_ptr // expected-error {{expected '(' after 'use_device_ptr'}}
+ {}
+#pragma omp target data map(i) use_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target data map(i) use_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(i) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(j) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(z) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(aa) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(e) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(g) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k,i,j) // expected-error2 {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(d) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(da) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+ return 0;
+}
+
+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;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+#pragma omp target data map(i) use_device_ptr // expected-error {{expected '(' after 'use_device_ptr'}}
+ {}
+#pragma omp target data map(i) use_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target data map(i) use_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target data map(i) use_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(i) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(j) // expected-error {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(z) // OK
+ {}
+#pragma omp target data map(i) use_device_ptr(aa) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(e) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(g) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(k,i,j) // expected-error2 {{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(d) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+#pragma omp target data map(i) use_device_ptr(da) // expected-error{{expected pointer or reference to pointer in 'use_device_ptr' clause}}
+ {}
+ return tmain<int, 3>(argc); // expected-note {{in instantiation of function template specialization 'tmain<int, 3>' requested here}}
+}
Modified: cfe/trunk/tools/libclang/CIndex.cpp
URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/tools/libclang/CIndex.cpp?rev=275271&r1=275270&r2=275271&view=diff
==============================================================================
--- cfe/trunk/tools/libclang/CIndex.cpp (original)
+++ cfe/trunk/tools/libclang/CIndex.cpp Wed Jul 13 10:37:16 2016
@@ -2275,6 +2275,9 @@ void OMPClauseEnqueue::VisitOMPToClause(
void OMPClauseEnqueue::VisitOMPFromClause(const OMPFromClause *C) {
VisitOMPClauseList(C);
}
+void OMPClauseEnqueue::VisitOMPUseDevicePtrClause(const OMPUseDevicePtrClause *C) {
+ VisitOMPClauseList(C);
+}
}
void EnqueueVisitor::EnqueueChildren(const OMPClause *S) {
More information about the cfe-commits
mailing list