Author: Alexey Bataev Date: 2019-10-29T10:31:24-04:00 New Revision: c09c0651a43b75044dc99e7c69acbcfaffd08aa2
URL: https://github.com/llvm/llvm-project/commit/c09c0651a43b75044dc99e7c69acbcfaffd08aa2 DIFF: https://github.com/llvm/llvm-project/commit/c09c0651a43b75044dc99e7c69acbcfaffd08aa2.diff LOG: [OPENMP]Fix PR43772: No warning in non-combined target regions. Need to analyze inner target regions in case of implicit mapping of the data members when target region is created in one of the class member functions. Added: Modified: clang/lib/Sema/SemaOpenMP.cpp clang/test/OpenMP/distribute_simd_loop_messages.cpp clang/test/OpenMP/teams_distribute_codegen.cpp clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp clang/test/OpenMP/teams_distribute_simd_codegen.cpp Removed: ################################################################################ diff --git a/clang/lib/Sema/SemaOpenMP.cpp b/clang/lib/Sema/SemaOpenMP.cpp index 8b1fca89dee8..b3b8fd655f14 100644 --- a/clang/lib/Sema/SemaOpenMP.cpp +++ b/clang/lib/Sema/SemaOpenMP.cpp @@ -2768,6 +2768,7 @@ class DSAAttrChecker final : public StmtVisitor<DSAAttrChecker, void> { DSAStackTy *Stack; Sema &SemaRef; bool ErrorFound = false; + bool TryCaptureCXXThisMembers = false; CapturedStmt *CS = nullptr; llvm::SmallVector<Expr *, 4> ImplicitFirstprivate; llvm::SmallVector<Expr *, 4> ImplicitMap; @@ -2779,12 +2780,26 @@ class DSAAttrChecker final : public StmtVisitor<DSAAttrChecker, void> { if (!S->hasAssociatedStmt() || !S->getAssociatedStmt()) return; visitSubCaptures(S->getInnermostCapturedStmt()); + // Try to capture inner this->member references to generate correct mappings + // and diagnostics. + if (TryCaptureCXXThisMembers || + (isOpenMPTargetExecutionDirective(Stack->getCurrentDirective()) && + llvm::any_of(S->getInnermostCapturedStmt()->captures(), + [](const CapturedStmt::Capture &C) { + return C.capturesThis(); + }))) { + bool SavedTryCaptureCXXThisMembers = TryCaptureCXXThisMembers; + TryCaptureCXXThisMembers = true; + Visit(S->getInnermostCapturedStmt()->getCapturedStmt()); + TryCaptureCXXThisMembers = SavedTryCaptureCXXThisMembers; + } } public: void VisitDeclRefExpr(DeclRefExpr *E) { - if (E->isTypeDependent() || E->isValueDependent() || - E->containsUnexpandedParameterPack() || E->isInstantiationDependent()) + if (TryCaptureCXXThisMembers || E->isTypeDependent() || + E->isValueDependent() || E->containsUnexpandedParameterPack() || + E->isInstantiationDependent()) return; if (auto *VD = dyn_cast<VarDecl>(E->getDecl())) { // Check the datasharing rules for the expressions in the clauses. @@ -3018,7 +3033,7 @@ class DSAAttrChecker final : public StmtVisitor<DSAAttrChecker, void> { })) { Visit(E->getBase()); } - } else { + } else if (!TryCaptureCXXThisMembers) { Visit(E->getBase()); } } diff --git a/clang/test/OpenMP/distribute_simd_loop_messages.cpp b/clang/test/OpenMP/distribute_simd_loop_messages.cpp index 423e5f1b1166..183fdfd90cfb 100644 --- a/clang/test/OpenMP/distribute_simd_loop_messages.cpp +++ b/clang/test/OpenMP/distribute_simd_loop_messages.cpp @@ -4,6 +4,22 @@ // RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -x c++ -std=c++11 -fexceptions -fcxx-exceptions -verify=expected,omp4 %s -Wuninitialized // RUN: %clang_cc1 -fsyntax-only -fopenmp-simd -fopenmp-version=50 -x c++ -std=c++11 -fexceptions -fcxx-exceptions -verify=expected,omp5 %s -Wuninitialized +class S5 { + int a; + S5() : a(0) {} + +public: + S5(int v) : a(v) {} + S5 &operator=(S5 &s) { +#pragma omp target +#pragma omp teams +#pragma omp distribute simd + for (int k = 0; k < s.a; ++k) // expected-warning {{Non-trivial type 'S5' is mapped, only trivial types are guaranteed to be mapped correctly}} + ++s.a; + return *this; + } +}; + static int sii; // expected-note@+1 {{defined as threadprivate or thread local}} #pragma omp threadprivate(sii) diff --git a/clang/test/OpenMP/teams_distribute_codegen.cpp b/clang/test/OpenMP/teams_distribute_codegen.cpp index 0065c6cbd51c..151fa2da49f9 100644 --- a/clang/test/OpenMP/teams_distribute_codegen.cpp +++ b/clang/test/OpenMP/teams_distribute_codegen.cpp @@ -158,7 +158,7 @@ struct SS{ // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( int foo(void) { - // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 1, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 0) + // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{.+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 0) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) #pragma omp target #pragma omp teams distribute @@ -174,7 +174,7 @@ struct SS{ // CK3: define internal void @[[OUTL1]]({{.+}}) // CK3: call void @__kmpc_for_static_init_4( // CK3: call void @__kmpc_for_static_fini( - // CK3: ret void + // CK3: ret void return a[0]; } diff --git a/clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp b/clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp index bc42f740ad74..43d3500d196f 100644 --- a/clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp +++ b/clang/test/OpenMP/teams_distribute_parallel_for_codegen.cpp @@ -161,7 +161,7 @@ struct SS{ // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( int foo(void) { - // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 1, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 0) + // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{.+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 0) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) #pragma omp target #pragma omp teams distribute parallel for diff --git a/clang/test/OpenMP/teams_distribute_simd_codegen.cpp b/clang/test/OpenMP/teams_distribute_simd_codegen.cpp index ff2a752cfa34..0f68328e67a6 100644 --- a/clang/test/OpenMP/teams_distribute_simd_codegen.cpp +++ b/clang/test/OpenMP/teams_distribute_simd_codegen.cpp @@ -165,7 +165,7 @@ struct SS{ // CK3: define {{.*}}i32 @{{.+}}foo{{.+}}( int foo(void) { - // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 1, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* {{.+}}@{{[^,]+}}, i32 0, i32 0), i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 1) + // CK3: call i32 @__tgt_target_teams(i64 -1, i8* @{{[^,]+}}, i32 2, i8** %{{[^,]+}}, i8** %{{[^,]+}}, i{{64|32}}* %{{.+}}, i64* {{.+}}@{{[^,]+}}, i32 0, i32 0), i32 0, i32 1) // CK3: call void @[[OFFL1:.+]]([[SSI]]* %{{.+}}) #pragma omp target #pragma omp teams distribute simd _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
