From 06e5f9a555256ed65aa2ed37d0d5e8ae47479d90 Mon Sep 17 00:00:00 2001 From: Alexey Bataev Date: Fri, 4 Aug 2017 19:10:54 +0000 Subject: [PATCH] [OPENMP] Unify generation of outlined function calls. git-svn-id: https://llvm.org/svn/llvm-project/cfe/trunk@310098 91177308-0d34-0410-b5e6-96231b3b80d8 --- lib/CodeGen/CGOpenMPRuntime.cpp | 20 ++++++++++++++++---- lib/CodeGen/CGOpenMPRuntime.h | 6 ++++++ lib/CodeGen/CGOpenMPRuntimeNVPTX.cpp | 10 +++++----- lib/CodeGen/CGStmtOpenMP.cpp | 7 ++++--- test/OpenMP/ordered_codegen.cpp | 4 ++-- 5 files changed, 33 insertions(+), 14 deletions(-) diff --git a/lib/CodeGen/CGOpenMPRuntime.cpp b/lib/CodeGen/CGOpenMPRuntime.cpp index d488bd4b30..80d90dd5f9 100644 --- a/lib/CodeGen/CGOpenMPRuntime.cpp +++ b/lib/CodeGen/CGOpenMPRuntime.cpp @@ -2447,7 +2447,7 @@ void CGOpenMPRuntime::emitParallelCall(CodeGenFunction &CGF, SourceLocation Loc, OutlinedFnArgs.push_back(ThreadIDAddr.getPointer()); OutlinedFnArgs.push_back(ZeroAddr.getPointer()); OutlinedFnArgs.append(CapturedVars.begin(), CapturedVars.end()); - CGF.EmitCallOrInvoke(OutlinedFn, OutlinedFnArgs); + RT.emitOutlinedFunctionCall(CGF, OutlinedFn, OutlinedFnArgs); // __kmpc_end_serialized_parallel(&Loc, GTid); llvm::Value *EndArgs[] = {RT.emitUpdateLocation(CGF, Loc), ThreadID}; @@ -3859,7 +3859,7 @@ emitProxyTaskFunction(CodeGenModule &CGM, SourceLocation Loc, } CallArgs.push_back(SharedsParam); - CGF.EmitCallOrInvoke(TaskFunction, CallArgs); + CGM.getOpenMPRuntime().emitOutlinedFunctionCall(CGF, TaskFunction, CallArgs); CGF.EmitStoreThroughLValue( RValue::get(CGF.Builder.getInt32(/*C=*/0)), CGF.MakeAddrLValue(CGF.ReturnValue, KmpInt32Ty)); @@ -4550,7 +4550,8 @@ void CGOpenMPRuntime::emitTaskCall(CodeGenFunction &CGF, SourceLocation Loc, CodeGenFunction &CGF, PrePostActionTy &Action) { Action.Enter(CGF); llvm::Value *OutlinedFnArgs[] = {ThreadID, NewTaskNewTaskTTy}; - CGF.EmitCallOrInvoke(TaskEntry, OutlinedFnArgs); + CGF.CGM.getOpenMPRuntime().emitOutlinedFunctionCall(CGF, TaskEntry, + OutlinedFnArgs); }; // Build void __kmpc_omp_task_begin_if0(ident_t *, kmp_int32 gtid, @@ -7034,7 +7035,7 @@ void CGOpenMPRuntime::emitTargetCall(CodeGenFunction &CGF, CGF.Builder.CreateCondBr(Failed, OffloadFailedBlock, OffloadContBlock); CGF.EmitBlock(OffloadFailedBlock); - CGF.Builder.CreateCall(OutlinedFn, KernelArgs); + emitOutlinedFunctionCall(CGF, OutlinedFn, KernelArgs); CGF.EmitBranch(OffloadContBlock); CGF.EmitBlock(OffloadContBlock, /*IsFinished=*/true); @@ -7754,3 +7755,14 @@ void CGOpenMPRuntime::emitDoacrossOrdered(CodeGenFunction &CGF, CGF.EmitRuntimeCall(RTLFn, Args); } +void CGOpenMPRuntime::emitOutlinedFunctionCall( + CodeGenFunction &CGF, llvm::Value *OutlinedFn, + ArrayRef Args) const { + if (auto *Fn = dyn_cast(OutlinedFn)) { + if (Fn->doesNotThrow()) { + CGF.EmitNounwindRuntimeCall(OutlinedFn, Args); + return; + } + } + CGF.EmitRuntimeCall(OutlinedFn, Args); +} diff --git a/lib/CodeGen/CGOpenMPRuntime.h b/lib/CodeGen/CGOpenMPRuntime.h index 5dcf999bea..809d9fa0a9 100644 --- a/lib/CodeGen/CGOpenMPRuntime.h +++ b/lib/CodeGen/CGOpenMPRuntime.h @@ -1324,6 +1324,12 @@ public: /// \param C 'depend' clause with 'sink|source' dependency kind. virtual void emitDoacrossOrdered(CodeGenFunction &CGF, const OMPDependClause *C); + + /// Emits call of the outlined function with the provided arguments, + /// translating these arguments to correct target-specific arguments. + virtual void + emitOutlinedFunctionCall(CodeGenFunction &CGF, llvm::Value *OutlinedFn, + ArrayRef Args = llvm::None) const; }; } // namespace CodeGen diff --git a/lib/CodeGen/CGOpenMPRuntimeNVPTX.cpp b/lib/CodeGen/CGOpenMPRuntimeNVPTX.cpp index 3ced05d08a..d6a8257796 100644 --- a/lib/CodeGen/CGOpenMPRuntimeNVPTX.cpp +++ b/lib/CodeGen/CGOpenMPRuntimeNVPTX.cpp @@ -345,7 +345,7 @@ void CGOpenMPRuntimeNVPTX::emitGenericEntryHeader(CodeGenFunction &CGF, Bld.CreateCondBr(IsWorker, WorkerBB, MasterCheckBB); CGF.EmitBlock(WorkerBB); - CGF.EmitCallOrInvoke(WST.WorkerFn, llvm::None); + emitOutlinedFunctionCall(CGF, WST.WorkerFn); CGF.EmitBranch(EST.ExitBB); CGF.EmitBlock(MasterCheckBB); @@ -555,7 +555,7 @@ void CGOpenMPRuntimeNVPTX::emitWorkerLoop(CodeGenFunction &CGF, CGF.CreateDefaultAlignTempAlloca(CGF.Int32Ty, /*Name=*/".zero.addr"); CGF.InitTempAlloca(ZeroAddr, CGF.Builder.getInt32(/*C=*/0)); llvm::Value *FnArgs[] = {ZeroAddr.getPointer(), ZeroAddr.getPointer()}; - CGF.EmitCallOrInvoke(Fn, FnArgs); + emitOutlinedFunctionCall(CGF, Fn, FnArgs); // Go to end of parallel region. CGF.EmitBranch(TerminateBB); @@ -883,7 +883,7 @@ void CGOpenMPRuntimeNVPTX::emitTeamsCall(CodeGenFunction &CGF, OutlinedFnArgs.push_back(ZeroAddr.getPointer()); OutlinedFnArgs.push_back(ZeroAddr.getPointer()); OutlinedFnArgs.append(CapturedVars.begin(), CapturedVars.end()); - CGF.EmitCallOrInvoke(OutlinedFn, OutlinedFnArgs); + emitOutlinedFunctionCall(CGF, OutlinedFn, OutlinedFnArgs); } void CGOpenMPRuntimeNVPTX::emitParallelCall( @@ -944,7 +944,7 @@ void CGOpenMPRuntimeNVPTX::emitGenericParallelCall( OutlinedFnArgs.push_back( llvm::ConstantPointerNull::get(CGM.Int32Ty->getPointerTo())); OutlinedFnArgs.append(CapturedVars.begin(), CapturedVars.end()); - CGF.EmitCallOrInvoke(Fn, OutlinedFnArgs); + emitOutlinedFunctionCall(CGF, Fn, OutlinedFnArgs); }; RegionCodeGenTy RCG(CodeGen); @@ -980,7 +980,7 @@ void CGOpenMPRuntimeNVPTX::emitSpmdParallelCall( OutlinedFnArgs.push_back( llvm::ConstantPointerNull::get(CGM.Int32Ty->getPointerTo())); OutlinedFnArgs.append(CapturedVars.begin(), CapturedVars.end()); - CGF.EmitCallOrInvoke(OutlinedFn, OutlinedFnArgs); + emitOutlinedFunctionCall(CGF, OutlinedFn, OutlinedFnArgs); } /// This function creates calls to one of two shuffle functions to copy diff --git a/lib/CodeGen/CGStmtOpenMP.cpp b/lib/CodeGen/CGStmtOpenMP.cpp index 57259a9912..436a04fbeb 100644 --- a/lib/CodeGen/CGStmtOpenMP.cpp +++ b/lib/CodeGen/CGStmtOpenMP.cpp @@ -328,7 +328,7 @@ static std::pair emitOutlinedFunctionPrologue( FO.FunctionName, &CGM.getModule()); CGM.SetInternalFunctionAttributes(CD, F, FuncInfo); if (CD->isNothrow()) - F->addFnAttr(llvm::Attribute::NoUnwind); + F->setDoesNotThrow(); // Generate the function. CGF.StartFunction(CD, Ctx.VoidTy, F, FuncInfo, Args, CD->getLocation(), @@ -482,7 +482,7 @@ CodeGenFunction::GenerateOpenMPCapturedStmtFunction(const CapturedStmt &S) { } CallArgs.emplace_back(CallArg); } - WrapperCGF.Builder.CreateCall(F, CallArgs); + CGM.getOpenMPRuntime().emitOutlinedFunctionCall(WrapperCGF, F, CallArgs); WrapperCGF.FinishFunction(); return WrapperF; } @@ -3151,7 +3151,8 @@ void CodeGenFunction::EmitOMPOrderedDirective(const OMPOrderedDirective &S) { llvm::SmallVector CapturedVars; CGF.GenerateOpenMPCapturedVars(*CS, CapturedVars); auto *OutlinedFn = emitOutlinedOrderedFunction(CGM, CS); - CGF.EmitNounwindRuntimeCall(OutlinedFn, CapturedVars); + CGM.getOpenMPRuntime().emitOutlinedFunctionCall(CGF, OutlinedFn, + CapturedVars); } else { Action.Enter(CGF); CGF.EmitStmt( diff --git a/test/OpenMP/ordered_codegen.cpp b/test/OpenMP/ordered_codegen.cpp index bd7c07051b..af6a872243 100644 --- a/test/OpenMP/ordered_codegen.cpp +++ b/test/OpenMP/ordered_codegen.cpp @@ -216,7 +216,7 @@ float f[10]; // CHECK-LABEL: foo_simd void foo_simd(int low, int up) { // CHECK: store float 0.000000e+00, float* %{{.+}}, align {{[0-9]+}}, !llvm.mem.parallel_loop_access ! - // CHECK-NEXT: call void [[CAP_FUNC:@.+]](i32* %{{.+}}) #{{[0-9]+}}, !llvm.mem.parallel_loop_access ! + // CHECK-NEXT: call void [[CAP_FUNC:@.+]](i32* %{{.+}}), !llvm.mem.parallel_loop_access ! #pragma omp simd for (int i = low; i < up; ++i) { f[i] = 0.0; @@ -224,7 +224,7 @@ void foo_simd(int low, int up) { f[i] = 1.0; } // CHECK: store float 0.000000e+00, float* %{{.+}}, align {{[0-9]+}} - // CHECK-NEXT: call void [[CAP_FUNC:@.+]](i32* %{{.+}}) #{{[0-9]+}} + // CHECK-NEXT: call void [[CAP_FUNC:@.+]](i32* %{{.+}}) #pragma omp for simd ordered for (int i = low; i < up; ++i) { f[i] = 0.0; -- 2.40.0