return T->getClauseKind() == OMPC_use_device_ptr;
}
};
+
+/// This represents clause 'is_device_ptr' in the '#pragma omp ...'
+/// directives.
+///
+/// \code
+/// #pragma omp target is_device_ptr(a,b)
+/// \endcode
+/// In this example directive '#pragma omp target' has clause
+/// 'is_device_ptr' with the variables 'a' and 'b'.
+///
+class OMPIsDevicePtrClause final
+ : public OMPVarListClause<OMPIsDevicePtrClause>,
+ private llvm::TrailingObjects<OMPIsDevicePtrClause, 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.
+ ///
+ OMPIsDevicePtrClause(SourceLocation StartLoc, SourceLocation LParenLoc,
+ SourceLocation EndLoc, unsigned N)
+ : OMPVarListClause<OMPIsDevicePtrClause>(OMPC_is_device_ptr, StartLoc,
+ LParenLoc, EndLoc, N) {}
+
+ /// Build an empty clause.
+ ///
+ /// \param N Number of variables.
+ ///
+ explicit OMPIsDevicePtrClause(unsigned N)
+ : OMPVarListClause<OMPIsDevicePtrClause>(
+ OMPC_is_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 OMPIsDevicePtrClause *
+ 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 OMPIsDevicePtrClause *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_is_device_ptr;
+ }
+};
} // end namespace clang
#endif // LLVM_CLANG_AST_OPENMPCLAUSE_H
return true;
}
+template <typename Derived>
+bool RecursiveASTVisitor<Derived>::VisitOMPIsDevicePtrClause(
+ OMPIsDevicePtrClause *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
"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 err_omp_argument_type_isdeviceptr : Error <
+ "expected pointer, array, reference to pointer, or reference to array in 'is_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>;
OPENMP_CLAUSE(to, OMPToClause)
OPENMP_CLAUSE(from, OMPFromClause)
OPENMP_CLAUSE(use_device_ptr, OMPUseDevicePtrClause)
+OPENMP_CLAUSE(is_device_ptr, OMPIsDevicePtrClause)
// Clauses allowed for OpenMP directive 'parallel'.
OPENMP_PARALLEL_CLAUSE(if)
OPENMP_ATOMIC_CLAUSE(seq_cst)
// Clauses allowed for OpenMP directive 'target'.
-// TODO More clauses for 'target' directive.
OPENMP_TARGET_CLAUSE(if)
OPENMP_TARGET_CLAUSE(device)
OPENMP_TARGET_CLAUSE(map)
OPENMP_TARGET_CLAUSE(depend)
OPENMP_TARGET_CLAUSE(defaultmap)
OPENMP_TARGET_CLAUSE(firstprivate)
+OPENMP_TARGET_CLAUSE(is_device_ptr)
// Clauses allowed for OpenMP directive 'target data'.
// TODO More clauses for 'target data' directive.
SourceLocation StartLoc,
SourceLocation LParenLoc,
SourceLocation EndLoc);
+ /// Called on well-formed 'is_device_ptr' clause.
+ OMPClause *ActOnOpenMPIsDevicePtrClause(ArrayRef<Expr *> VarList,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc);
/// \brief The kind of conversion being performed.
enum CheckedConversionKind {
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
break;
}
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
break;
}
void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(N));
return new (Mem) OMPUseDevicePtrClause(N);
}
+
+OMPIsDevicePtrClause *OMPIsDevicePtrClause::Create(const ASTContext &C,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc,
+ ArrayRef<Expr *> VL) {
+ void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(VL.size()));
+ OMPIsDevicePtrClause *Clause =
+ new (Mem) OMPIsDevicePtrClause(StartLoc, LParenLoc, EndLoc, VL.size());
+ Clause->setVarRefs(VL);
+ return Clause;
+}
+
+OMPIsDevicePtrClause *OMPIsDevicePtrClause::CreateEmpty(const ASTContext &C,
+ unsigned N) {
+ void *Mem = C.Allocate(totalSizeToAlloc<Expr *>(N));
+ return new (Mem) OMPIsDevicePtrClause(N);
+}
OS << ")";
}
}
+
+void OMPClausePrinter::VisitOMPIsDevicePtrClause(OMPIsDevicePtrClause *Node) {
+ if (!Node->varlist_empty()) {
+ OS << "is_device_ptr";
+ VisitOMPClauseList(Node, '(');
+ OS << ")";
+ }
+}
}
//===----------------------------------------------------------------------===//
const OMPUseDevicePtrClause *C) {
VisitOMPClauseList(C);
}
+void OMPClauseProfiler::VisitOMPIsDevicePtrClause(
+ const OMPIsDevicePtrClause *C) {
+ VisitOMPClauseList(C);
+}
}
void
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
break;
}
llvm_unreachable("Invalid OpenMP simple clause kind");
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
break;
}
llvm_unreachable("Invalid OpenMP simple clause kind");
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
llvm_unreachable("Clause is not allowed in 'omp atomic'.");
}
}
/// simdlen-clause | threads-clause | simd-clause | num_teams-clause |
/// thread_limit-clause | priority-clause | grainsize-clause |
/// nogroup-clause | num_tasks-clause | hint-clause | to-clause |
-/// from-clause
+/// from-clause | is_device_ptr-clause
///
OMPClause *Parser::ParseOpenMPClause(OpenMPDirectiveKind DKind,
OpenMPClauseKind CKind, bool FirstClause) {
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
Clause = ParseOpenMPVarListClause(DKind, CKind);
break;
case OMPC_unknown:
/// 'from' '(' list ')'
/// use_device_ptr-clause:
/// 'use_device_ptr' '(' list ')'
+/// is_device_ptr-clause:
+/// 'is_device_ptr' '(' list ')'
///
/// For 'linear' clause linear-list may have the following forms:
/// list
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
case OMPC_to:
case OMPC_from:
case OMPC_use_device_ptr:
+ case OMPC_is_device_ptr:
llvm_unreachable("Clause is not allowed.");
}
return Res;
case OMPC_use_device_ptr:
Res = ActOnOpenMPUseDevicePtrClause(VarList, StartLoc, LParenLoc, EndLoc);
break;
+ case OMPC_is_device_ptr:
+ Res = ActOnOpenMPIsDevicePtrClause(VarList, StartLoc, LParenLoc, EndLoc);
+ break;
case OMPC_if:
case OMPC_final:
case OMPC_num_threads:
return OMPUseDevicePtrClause::Create(Context, StartLoc, LParenLoc, EndLoc,
Vars);
}
+
+OMPClause *Sema::ActOnOpenMPIsDevicePtrClause(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 array or reference to pointer or array
+ if (!Type.getNonReferenceType()->isPointerType() &&
+ !Type.getNonReferenceType()->isArrayType()) {
+ Diag(ELoc, diag::err_omp_argument_type_isdeviceptr)
+ << 0 << RefExpr->getSourceRange();
+ continue;
+ }
+ Vars.push_back(RefExpr->IgnoreParens());
+ }
+
+ if (Vars.empty())
+ return nullptr;
+
+ return OMPIsDevicePtrClause::Create(Context, StartLoc, LParenLoc, EndLoc,
+ Vars);
+}
EndLoc);
}
+ /// Build a new OpenMP 'is_device_ptr' clause.
+ ///
+ /// By default, performs semantic analysis to build the new OpenMP clause.
+ /// Subclasses may override this routine to provide different behavior.
+ OMPClause *RebuildOMPIsDevicePtrClause(ArrayRef<Expr *> VarList,
+ SourceLocation StartLoc,
+ SourceLocation LParenLoc,
+ SourceLocation EndLoc) {
+ return getSema().ActOnOpenMPIsDevicePtrClause(VarList, StartLoc, LParenLoc,
+ EndLoc);
+ }
+
/// \brief Rebuild the operand to an Objective-C \@synchronized statement.
///
/// By default, performs semantic analysis to build the new statement.
Vars, C->getLocStart(), C->getLParenLoc(), C->getLocEnd());
}
+template <typename Derived>
+OMPClause *
+TreeTransform<Derived>::TransformOMPIsDevicePtrClause(OMPIsDevicePtrClause *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().RebuildOMPIsDevicePtrClause(
+ Vars, C->getLocStart(), C->getLParenLoc(), C->getLocEnd());
+}
+
//===----------------------------------------------------------------------===//
// Expression transformation
//===----------------------------------------------------------------------===//
case OMPC_use_device_ptr:
C = OMPUseDevicePtrClause::CreateEmpty(Context, Record[Idx++]);
break;
+ case OMPC_is_device_ptr:
+ C = OMPIsDevicePtrClause::CreateEmpty(Context, Record[Idx++]);
+ break;
}
Visit(C);
C->setLocStart(Reader->ReadSourceLocation(Record, Idx));
Vars.clear();
}
+void OMPClauseReader::VisitOMPIsDevicePtrClause(OMPIsDevicePtrClause *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.
//===----------------------------------------------------------------------===//
}
}
+void OMPClauseWriter::VisitOMPIsDevicePtrClause(OMPIsDevicePtrClause *C) {
+ Record.push_back(C->varlist_size());
+ Record.AddSourceLocation(C->getLParenLoc());
+ for (auto *VE : C->varlists()) {
+ Record.AddStmt(VE);
+ }
+}
+
//===----------------------------------------------------------------------===//
// OpenMP Directives.
//===----------------------------------------------------------------------===//
--- /dev/null
+// RUN: %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;
+};
+typedef int arr[10];
+typedef ST STarr[10];
+struct SA {
+ const int da[5] = { 0 };
+ ST g[10];
+ STarr &rg = g;
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+ arr &raa = aa;
+ void func(int arg) {
+#pragma omp target is_device_ptr(k)
+ {}
+#pragma omp target is_device_ptr(z)
+ {}
+#pragma omp target is_device_ptr(aa) // OK
+ {}
+#pragma omp target is_device_ptr(raa) // OK
+ {}
+#pragma omp target is_device_ptr(g) // OK
+ {}
+#pragma omp target is_device_ptr(rg) // OK
+ {}
+#pragma omp target is_device_ptr(da) // OK
+ {}
+ return;
+ }
+};
+// CHECK: struct SA
+// CHECK-NEXT: const int da[5] = {0};
+// CHECK-NEXT: ST g[10];
+// CHECK-NEXT: STarr &rg = this->g;
+// CHECK-NEXT: int i;
+// CHECK-NEXT: int &j = this->i;
+// CHECK-NEXT: int *k = &this->j;
+// CHECK-NEXT: int *&z = this->k;
+// CHECK-NEXT: int aa[10];
+// CHECK-NEXT: arr &raa = this->aa;
+// CHECK-NEXT: func(
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->k)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->z)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->aa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->raa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->g)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->rg)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(this->da)
+
+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 struct {
+ int a;
+} S6;
+
+template <typename T>
+T tmain(T argc) {
+ const T da[5] = { 0 };
+ S6 h[10];
+ auto &rh = h;
+ T i;
+ T &j = i;
+ T *k = &j;
+ T *&z = k;
+ T aa[10];
+ auto &raa = aa;
+#pragma omp target is_device_ptr(k)
+ {}
+#pragma omp target is_device_ptr(z)
+ {}
+#pragma omp target is_device_ptr(aa)
+ {}
+#pragma omp target is_device_ptr(raa)
+ {}
+#pragma omp target is_device_ptr(h)
+ {}
+#pragma omp target is_device_ptr(rh)
+ {}
+#pragma omp target is_device_ptr(da)
+ {}
+ return 0;
+}
+
+// CHECK: template <typename T = int> int tmain(int argc) {
+// CHECK-NEXT: const int da[5] = {0};
+// CHECK-NEXT: S6 h[10];
+// CHECK-NEXT: auto &rh = h;
+// CHECK-NEXT: int i;
+// CHECK-NEXT: int &j = i;
+// CHECK-NEXT: int *k = &j;
+// CHECK-NEXT: int *&z = k;
+// CHECK-NEXT: int aa[10];
+// CHECK-NEXT: auto &raa = aa;
+// CHECK-NEXT: #pragma omp target is_device_ptr(k)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(z)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(aa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(raa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(h)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(rh)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(da)
+
+// CHECK: template <typename T = int *> int *tmain(int *argc) {
+// CHECK-NEXT: int *const da[5] = {0};
+// CHECK-NEXT: S6 h[10];
+// CHECK-NEXT: auto &rh = h;
+// CHECK-NEXT: int *i;
+// CHECK-NEXT: int *&j = i;
+// CHECK-NEXT: int **k = &j;
+// CHECK-NEXT: int **&z = k;
+// CHECK-NEXT: int *aa[10];
+// CHECK-NEXT: auto &raa = aa;
+// CHECK-NEXT: #pragma omp target is_device_ptr(k)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(z)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(aa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(raa)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(h)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(rh)
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+// CHECK-NEXT: #pragma omp target is_device_ptr(da)
+
+// CHECK-LABEL: int main(int argc, char **argv) {
+int main(int argc, char **argv) {
+ const int da[5] = { 0 };
+ S6 h[10];
+ auto &rh = h;
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+ auto &raa = aa;
+// CHECK-NEXT: const int da[5] = {0};
+// CHECK-NEXT: S6 h[10];
+// CHECK-NEXT: auto &rh = h;
+// CHECK-NEXT: int i;
+// CHECK-NEXT: int &j = i;
+// CHECK-NEXT: int *k = &j;
+// CHECK-NEXT: int *&z = k;
+// CHECK-NEXT: int aa[10];
+// CHECK-NEXT: auto &raa = aa;
+#pragma omp target is_device_ptr(k)
+// CHECK-NEXT: #pragma omp target is_device_ptr(k)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(z)
+// CHECK-NEXT: #pragma omp target is_device_ptr(z)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(aa)
+// CHECK-NEXT: #pragma omp target is_device_ptr(aa)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(raa)
+// CHECK-NEXT: #pragma omp target is_device_ptr(raa)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(h)
+// CHECK-NEXT: #pragma omp target is_device_ptr(h)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(rh)
+// CHECK-NEXT: #pragma omp target is_device_ptr(rh)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+#pragma omp target is_device_ptr(da)
+// CHECK-NEXT: #pragma omp target is_device_ptr(da)
+ {}
+// CHECK-NEXT: {
+// CHECK-NEXT: }
+ return tmain<int>(argc) + *tmain<int *>(&argc);
+}
+
+
+#endif
--- /dev/null
+// RUN: %clang_cc1 -std=c++11 -verify -fopenmp -ferror-limit 200 %s
+struct ST {
+ int *a;
+};
+typedef int arr[10];
+typedef ST STarr[10];
+struct SA {
+ const int d = 5;
+ const int da[5] = { 0 };
+ ST e;
+ ST g[10];
+ STarr &rg = g;
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+ arr &raa = aa;
+ void func(int arg) {
+#pragma omp target is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}}
+ {}
+#pragma omp target is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target is_device_ptr(arg // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(k) // OK
+ {}
+#pragma omp target is_device_ptr(z) // OK
+ {}
+#pragma omp target is_device_ptr(aa) // OK
+ {}
+#pragma omp target is_device_ptr(raa) // OK
+ {}
+#pragma omp target is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(g) // OK
+ {}
+#pragma omp target is_device_ptr(rg) // OK
+ {}
+#pragma omp target is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(da) // OK
+ {}
+ 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 struct {
+ int a;
+} S6;
+
+template <typename T, int I>
+T tmain(T argc) {
+ const T d = 5;
+ const T da[5] = { 0 };
+ S4 e(4);
+ S5 g(5);
+ S6 h[10];
+ auto &rh = h;
+ T i;
+ T &j = i;
+ T *k = &j;
+ T *&z = k;
+ T aa[10];
+ auto &raa = aa;
+#pragma omp target is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}}
+ {}
+#pragma omp target is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target is_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(k) // OK
+ {}
+#pragma omp target is_device_ptr(z) // OK
+ {}
+#pragma omp target is_device_ptr(aa) // OK
+ {}
+#pragma omp target is_device_ptr(raa) // OK
+ {}
+#pragma omp target is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(g) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(h) // OK
+ {}
+#pragma omp target is_device_ptr(rh) // OK
+ {}
+#pragma omp target is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(da) // OK
+ {}
+ return 0;
+}
+
+int main(int argc, char **argv) {
+ const int d = 5;
+ const int da[5] = { 0 };
+ S4 e(4);
+ S5 g(5);
+ S6 h[10];
+ auto &rh = h;
+ int i;
+ int &j = i;
+ int *k = &j;
+ int *&z = k;
+ int aa[10];
+ auto &raa = aa;
+#pragma omp target is_device_ptr // expected-error {{expected '(' after 'is_device_ptr'}}
+ {}
+#pragma omp target is_device_ptr( // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr() // expected-error {{expected expression}}
+ {}
+#pragma omp target is_device_ptr(alloc) // expected-error {{use of undeclared identifier 'alloc'}}
+ {}
+#pragma omp target is_device_ptr(argc // expected-error {{expected ')'}} expected-note {{to match this '('}} expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(i) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(j) // expected-error {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(k) // OK
+ {}
+#pragma omp target is_device_ptr(z) // OK
+ {}
+#pragma omp target is_device_ptr(aa) // OK
+ {}
+#pragma omp target is_device_ptr(raa) // OK
+ {}
+#pragma omp target is_device_ptr(e) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(g) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(h) // OK
+ {}
+#pragma omp target is_device_ptr(rh) // OK
+ {}
+#pragma omp target is_device_ptr(k,i,j) // expected-error2 {{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(d) // expected-error{{expected pointer, array, reference to pointer, or reference to array in 'is_device_ptr clause'}}
+ {}
+#pragma omp target is_device_ptr(da) // OK
+ {}
+ return tmain<int, 3>(argc); // expected-note {{in instantiation of function template specialization 'tmain<int, 3>' requested here}}
+}
void OMPClauseEnqueue::VisitOMPUseDevicePtrClause(const OMPUseDevicePtrClause *C) {
VisitOMPClauseList(C);
}
+void OMPClauseEnqueue::VisitOMPIsDevicePtrClause(const OMPIsDevicePtrClause *C) {
+ VisitOMPClauseList(C);
+}
}
void EnqueueVisitor::EnqueueChildren(const OMPClause *S) {
projects / clang / commitdiff