Author: abataev
Date: Tue Aug 22 10:54:52 2017
New Revision: 311479
URL: http://llvm.org/viewvc/llvm-project?rev=311479&view=rev
Log:
[OPENMP] Fix for PR34014: OpenMP 4.5: Target construct in static method
of class fails to map class static variable.
If the global variable is captured and it has several redeclarations,
sometimes it may lead to a compiler crash. Patch fixes this by working
only with canonical declarations.
Modified:
cfe/trunk/lib/CodeGen/CGExpr.cpp
cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp
cfe/trunk/lib/CodeGen/CodeGenFunction.h
cfe/trunk/lib/Sema/SemaExpr.cpp
cfe/trunk/test/OpenMP/target_map_codegen.cpp
Modified: cfe/trunk/lib/CodeGen/CGExpr.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/CodeGen/CGExpr.cpp?rev=311479&r1=311478&r2=311479&view=diff
==============================================================================
--- cfe/trunk/lib/CodeGen/CGExpr.cpp (original)
+++ cfe/trunk/lib/CodeGen/CGExpr.cpp Tue Aug 22 10:54:52 2017
@@ -2300,6 +2300,7 @@ LValue CodeGenFunction::EmitDeclRefLValu
// Check for captured variables.
if (E->refersToEnclosingVariableOrCapture()) {
+ VD = VD->getCanonicalDecl();
if (auto *FD = LambdaCaptureFields.lookup(VD))
return EmitCapturedFieldLValue(*this, FD, CXXABIThisValue);
else if (CapturedStmtInfo) {
Modified: cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp?rev=311479&r1=311478&r2=311479&view=diff
==============================================================================
--- cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp (original)
+++ cfe/trunk/lib/CodeGen/CGStmtOpenMP.cpp Tue Aug 22 10:54:52 2017
@@ -65,6 +65,8 @@ public:
for (auto &C : CS->captures()) {
if (C.capturesVariable() || C.capturesVariableByCopy()) {
auto *VD = C.getCapturedVar();
+ assert(VD == VD->getCanonicalDecl() &&
+ "Canonical decl must be captured.");
DeclRefExpr DRE(const_cast<VarDecl *>(VD),
isCapturedVar(CGF, VD) ||
(CGF.CapturedStmtInfo &&
Modified: cfe/trunk/lib/CodeGen/CodeGenFunction.h
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/CodeGen/CodeGenFunction.h?rev=311479&r1=311478&r2=311479&view=diff
==============================================================================
--- cfe/trunk/lib/CodeGen/CodeGenFunction.h (original)
+++ cfe/trunk/lib/CodeGen/CodeGenFunction.h Tue Aug 22 10:54:52 2017
@@ -709,6 +709,7 @@ public:
llvm::function_ref<Address()> PrivateGen) {
assert(PerformCleanup && "adding private to dead scope");
+ LocalVD = LocalVD->getCanonicalDecl();
// Only save it once.
if (SavedLocals.count(LocalVD)) return false;
@@ -761,6 +762,7 @@ public:
/// Checks if the global variable is captured in current function.
bool isGlobalVarCaptured(const VarDecl *VD) const {
+ VD = VD->getCanonicalDecl();
return !VD->isLocalVarDeclOrParm() && CGF.LocalDeclMap.count(VD) > 0;
}
Modified: cfe/trunk/lib/Sema/SemaExpr.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Sema/SemaExpr.cpp?rev=311479&r1=311478&r2=311479&view=diff
==============================================================================
--- cfe/trunk/lib/Sema/SemaExpr.cpp (original)
+++ cfe/trunk/lib/Sema/SemaExpr.cpp Tue Aug 22 10:54:52 2017
@@ -14200,6 +14200,7 @@ bool Sema::tryCaptureVariable(
bool IsGlobal = !Var->hasLocalStorage();
if (IsGlobal && !(LangOpts.OpenMP && IsOpenMPCapturedDecl(Var)))
return true;
+ Var = Var->getCanonicalDecl();
// Walk up the stack to determine whether we can capture the variable,
// performing the "simple" checks that don't depend on type. We stop when
Modified: cfe/trunk/test/OpenMP/target_map_codegen.cpp
URL:
http://llvm.org/viewvc/llvm-project/cfe/trunk/test/OpenMP/target_map_codegen.cpp?rev=311479&r1=311478&r2=311479&view=diff
==============================================================================
--- cfe/trunk/test/OpenMP/target_map_codegen.cpp (original)
+++ cfe/trunk/test/OpenMP/target_map_codegen.cpp Tue Aug 22 10:54:52 2017
@@ -15,12 +15,30 @@
// RUN: %clang_cc1 -fopenmp -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 CK1 --check-prefix CK1-32
#ifdef CK1
+class B {
+public:
+ static double VAR;
+ B() {
+ }
+
+ static void modify(int &res) {
+#pragma omp target map(tofrom \
+ : res)
+ {
+ res = B::VAR;
+ }
+ }
+};
+double B::VAR = 1.0;
+
// CK1-DAG: [[SIZES:@.+]] = {{.+}}constant [1 x i[[sz:64|32]]] [i{{64|32}} 4]
// Map types: OMP_MAP_PRIVATE_VAL | OMP_MAP_IS_FIRST = 288
// CK1-DAG: [[TYPES:@.+]] = {{.+}}constant [1 x i32] [i32 288]
// CK1-LABEL: implicit_maps_integer
void implicit_maps_integer (int a){
+ // CK1: call void{{.*}}modify
+ B::modify(a);
int i = a;
// CK1-DAG: call i32 @__tgt_target(i32 {{.+}}, i8* {{.+}}, i32 1, i8**
[[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}},
{{.+}}[[TYPES]]{{.+}})
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
