From: Alexey Bataev <a.bataev@hotmail.com>
Date: Tue, 17 Oct 2017 16:47:34 +0000 (+0000)
Subject: [OPENMP] Fix capturing of boolean variables in debug mode.
X-Git-Url: https://granicus.if.org/sourcecode?a=commitdiff_plain;h=1b641bcb61ddab51b491e8875e3994604ac1ae17;p=clang

[OPENMP] Fix capturing of boolean variables in debug mode.

If the variables is boolean and we generating inner function with real
types, the codegen may crash because of not loading boolean value from
memory.

git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@316011 91177308-0d34-0410-b5e6-96231b3b80d8
---

diff --git a/lib/CodeGen/CGStmtOpenMP.cpp b/lib/CodeGen/CGStmtOpenMP.cpp
index 66d248d6ff..cdec3e35f9 100644
--- a/lib/CodeGen/CGStmtOpenMP.cpp
+++ b/lib/CodeGen/CGStmtOpenMP.cpp
@@ -501,9 +501,10 @@ CodeGenFunction::GenerateOpenMPCapturedStmtFunction(const CapturedStmt &S) {
     llvm::Value *CallArg;
     auto I = LocalAddrs.find(Arg);
     if (I != LocalAddrs.end()) {
-      LValue LV =
-          WrapperCGF.MakeAddrLValue(I->second.second, Arg->getType(),
-                                    AlignmentSource::Decl);
+      LValue LV = WrapperCGF.MakeAddrLValue(
+          I->second.second,
+          I->second.first ? I->second.first->getType() : Arg->getType(),
+          AlignmentSource::Decl);
       CallArg = WrapperCGF.EmitLoadOfScalar(LV, SourceLocation());
     } else {
       auto EI = VLASizes.find(Arg);
@@ -516,7 +517,7 @@ CodeGenFunction::GenerateOpenMPCapturedStmtFunction(const CapturedStmt &S) {
         CallArg = WrapperCGF.EmitLoadOfScalar(LV, SourceLocation());
       }
     }
-    CallArgs.emplace_back(CallArg);
+    CallArgs.emplace_back(WrapperCGF.EmitFromMemory(CallArg, Arg->getType()));
   }
   CGM.getOpenMPRuntime().emitOutlinedFunctionCall(WrapperCGF, S.getLocStart(),
                                                   F, CallArgs);
diff --git a/test/OpenMP/target_parallel_debug_codegen.cpp b/test/OpenMP/target_parallel_debug_codegen.cpp
index 39f714e547..cd19a8fdbc 100644
--- a/test/OpenMP/target_parallel_debug_codegen.cpp
+++ b/test/OpenMP/target_parallel_debug_codegen.cpp
@@ -11,7 +11,7 @@ int main() {
   int c[10][10][10];
 #pragma omp target parallel firstprivate(a, b) map(tofrom          \
                                                    : c) map(tofrom \
-                                                            : bb)
+                                                            : bb) if (a)
   {
     int &f = c[1][1][1];
     int &g = a;
@@ -54,7 +54,7 @@ int main() {
   return 0;
 }
 
-// CHECK: define internal void @__omp_offloading{{[^(]+}}([10 x [10 x [10 x i32]]] addrspace(1)* {{[^,]+}}, i32 {{[^,]+}}, [10 x [10 x i32]]* {{[^,]+}}, i8 addrspace(1)* noalias{{[^)]+}})
+// CHECK: define internal void @__omp_offloading{{[^(]+}}([10 x [10 x [10 x i32]]] addrspace(1)* {{[^,]+}}, i32 {{[^,]+}}, [10 x [10 x i32]]* {{[^,]+}}, i8 addrspace(1)* noalias{{[^,]+}}, i1 {{[^)]+}})
 // CHECK: addrspacecast [10 x [10 x [10 x i32]]] addrspace(1)* %{{.+}} to [10 x [10 x [10 x i32]]]*
 // CHECK: call void [[NONDEBUG_WRAPPER:.+]](i32* {{[^,]+}}, i32* {{[^,]+}}, [10 x [10 x [10 x i32]]]* {{[^,]+}}, i64 {{[^,]+}}, [10 x [10 x i32]]* {{[^,]+}}, i8* {{[^)]+}})