From fbd6d2c54e57a4968d29bb22742dd49759b3ecd0 Mon Sep 17 00:00:00 2001 From: Alexey Bataev Date: Wed, 12 Aug 2020 11:48:53 -0400 Subject: [PATCH] [OPENMP] Fix PR47063: crash when trying to get captured statetment. Need to call getRawStmt() function instead, when trying to get inner associated statement for the executable directive. Not all directives use captured statements. --- clang/lib/CodeGen/CGOpenMPRuntime.cpp | 3 +-- clang/test/OpenMP/target_codegen.cpp | 6 +++++- 2 files changed, 6 insertions(+), 3 deletions(-) diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp index 66d1b36dbb6a..c55403920d8f 100644 --- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp +++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp @@ -9801,8 +9801,7 @@ void CGOpenMPRuntime::scanForTargetRegionsFunctions(const Stmt *S, if (!E->hasAssociatedStmt() || !E->getAssociatedStmt()) return; - scanForTargetRegionsFunctions( - E->getInnermostCapturedStmt()->getCapturedStmt(), ParentName); + scanForTargetRegionsFunctions(E->getRawStmt(), ParentName); return; } diff --git a/clang/test/OpenMP/target_codegen.cpp b/clang/test/OpenMP/target_codegen.cpp index 9cec6bfa5a48..55b03aae00a5 100644 --- a/clang/test/OpenMP/target_codegen.cpp +++ b/clang/test/OpenMP/target_codegen.cpp @@ -712,6 +712,8 @@ int bar(int n){ // OMP45: [[BP:%.+]] = alloca [1 x i8*] // OMP45: [[P:%.+]] = alloca [1 x i8*] // OMP45: [[LOCAL_THIS1:%.+]] = load [[S2]]*, [[S2]]** [[LOCAL_THIS]] + +// OMP45: call void @__kmpc_critical( // OMP45: [[ARR_IDX:%.+]] = getelementptr inbounds [[S2]], [[S2]]* [[LOCAL_THIS1]], i[[SZ]] 0 // OMP45: [[ARR_IDX2:%.+]] = getelementptr inbounds [[S2]], [[S2]]* [[LOCAL_THIS1]], i[[SZ]] 0 @@ -731,6 +733,7 @@ int bar(int n){ // OMP45: call void [[HVT0:@.+]]([[S2]]* [[LOCAL_THIS1]]) // OMP45-NEXT: br label %[[END]] // OMP45: [[END]] +// OMP45: call void @__kmpc_end_critical( // Check that the offloading functions are emitted and that the arguments are // correct and loaded correctly for the target regions of the callees of bar(). @@ -826,7 +829,7 @@ int bar(int n){ // OMP50: call void [[HVT0:@.+]]([[S2]]* [[LOCAL_THIS1]]) // OMP50-NEXT: br label %[[END]] // OMP50: [[END]] - + void bar () { #define pragma_target _Pragma("omp target") pragma_target @@ -838,6 +841,7 @@ class S2 { public: void zee() { +#pragma omp critical #pragma omp target map(this[0]) a++; }