diff --git a/clang/lib/CodeGen/CGOpenMPRuntime.cpp b/clang/lib/CodeGen/CGOpenMPRuntime.cpp index 0d8b5c2fff85..3de5d3d84fe0 100644 --- a/clang/lib/CodeGen/CGOpenMPRuntime.cpp +++ b/clang/lib/CodeGen/CGOpenMPRuntime.cpp @@ -1293,6 +1293,13 @@ static llvm::Value *emitParallelOrTeamsOutlinedFunction( HasCancel = OPFD->hasCancel(); else if (auto *OPFD = dyn_cast(&D)) HasCancel = OPFD->hasCancel(); + else if (auto *OPFD = dyn_cast(&D)) + HasCancel = OPFD->hasCancel(); + else if (auto *OPFD = dyn_cast(&D)) + HasCancel = OPFD->hasCancel(); + else if (auto *OPFD = + dyn_cast(&D)) + HasCancel = OPFD->hasCancel(); CGOpenMPOutlinedRegionInfo CGInfo(*CS, ThreadIDVar, CodeGen, InnermostKind, HasCancel, OutlinedHelperName); CodeGenFunction::CGCapturedStmtRAII CapInfoRAII(CGF, &CGInfo); diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp index af9f7e2eff51..22a36eb0b555 100644 --- a/clang/lib/CodeGen/CGStmtOpenMP.cpp +++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp @@ -2039,8 +2039,7 @@ void CodeGenFunction::EmitOMPDistributeParallelForDirective( S.getDistInc()); }; OMPLexicalScope Scope(*this, S, /*AsInlined=*/true); - CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_distribute, CodeGen, - S.hasCancel()); + CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_distribute, CodeGen); } void CodeGenFunction::EmitOMPDistributeParallelForSimdDirective( @@ -3201,8 +3200,7 @@ void CodeGenFunction::EmitOMPDistributeDirective( CGF.EmitOMPDistributeLoop(S, emitOMPLoopBodyWithStopPoint, S.getInc()); }; OMPLexicalScope Scope(*this, S, /*AsInlined=*/true); - CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_distribute, CodeGen, - false); + CGM.getOpenMPRuntime().emitInlinedDirective(*this, OMPD_distribute, CodeGen); } static llvm::Function *emitOutlinedOrderedFunction(CodeGenModule &CGM, @@ -3915,8 +3913,8 @@ void CodeGenFunction::EmitOMPTeamsDistributeParallelForDirective( OMPPrivateScope PrivateScope(CGF); CGF.EmitOMPReductionClauseInit(S, PrivateScope); (void)PrivateScope.Privatize(); - CGF.CGM.getOpenMPRuntime().emitInlinedDirective( - CGF, OMPD_distribute, CodeGenDistribute, S.hasCancel()); + CGF.CGM.getOpenMPRuntime().emitInlinedDirective(CGF, OMPD_distribute, + CodeGenDistribute); CGF.EmitOMPReductionClauseFinal(S, /*ReductionKind=*/OMPD_teams); }; emitCommonOMPTeamsDirective(*this, S, OMPD_distribute_parallel_for, CodeGen); diff --git a/clang/test/OpenMP/target_parallel_for_codegen.cpp b/clang/test/OpenMP/target_parallel_for_codegen.cpp index 2b725fcbcdd5..55dd4d2a578d 100644 --- a/clang/test/OpenMP/target_parallel_for_codegen.cpp +++ b/clang/test/OpenMP/target_parallel_for_codegen.cpp @@ -106,6 +106,7 @@ int foo(int n) { #pragma omp target parallel for for (int i = 3; i < 32; i += 5) { #pragma omp cancel for +#pragma omp cancellation point for } // CHECK: call void [[HVT1:@.+]](i[[SZ]] {{[^,]+}}, i{{32|64}}{{[*]*}} {{[^)]+}}) @@ -325,6 +326,7 @@ int foo(int n) { // // CHECK: define internal {{.*}}void [[OMP_OUTLINED]](i32* noalias %.global_tid., i32* noalias %.bound_tid.) // CHECK: call i32 @__kmpc_cancel(%ident_t* @ +// CHECK: call i32 @__kmpc_cancellationpoint(%ident_t* @ // CHECK: ret void // CHECK: }