Index: include/clang/AST/Decl.h =================================================================== --- include/clang/AST/Decl.h +++ include/clang/AST/Decl.h @@ -3358,12 +3358,14 @@ private: /// \brief The number of parameters to the outlined function. unsigned NumParams; + /// \brief The position of context parameter in list of parameters. + unsigned ContextParam; /// \brief The body of the outlined function. - Stmt *Body; + llvm::PointerIntPair BodyAndNothrow; explicit CapturedDecl(DeclContext *DC, unsigned NumParams) : Decl(Captured, DC, SourceLocation()), DeclContext(Captured), - NumParams(NumParams), Body(0) { } + NumParams(NumParams), ContextParam(0), BodyAndNothrow(0, false) { } ImplicitParamDecl **getParams() const { return reinterpret_cast( @@ -3371,12 +3373,16 @@ } public: - static CapturedDecl *Create(ASTContext &C, DeclContext *DC, unsigned NumParams); + static CapturedDecl *Create(ASTContext &C, DeclContext *DC, + unsigned NumParams); static CapturedDecl *CreateDeserialized(ASTContext &C, unsigned ID, unsigned NumParams); - Stmt *getBody() const { return Body; } - void setBody(Stmt *B) { Body = B; } + Stmt *getBody() const { return BodyAndNothrow.getPointer(); } + void setBody(Stmt *B) { BodyAndNothrow.setPointer(B); } + + bool isNothrow() const { return BodyAndNothrow.getInt(); } + void setNothrow(bool Nothrow = true) { BodyAndNothrow.setInt(Nothrow); } unsigned getNumParams() const { return NumParams; } @@ -3390,8 +3396,16 @@ } /// \brief Retrieve the parameter containing captured variables. - ImplicitParamDecl *getContextParam() const { return getParam(0); } - void setContextParam(ImplicitParamDecl *P) { setParam(0, P); } + ImplicitParamDecl *getContextParam() const { + assert(ContextParam < NumParams); + return getParam(ContextParam); + } + void setContextParam(unsigned i, ImplicitParamDecl *P) { + assert(i < NumParams); + ContextParam = i; + setParam(i, P); + } + unsigned getContextParamPosition() const { return ContextParam; } typedef ImplicitParamDecl **param_iterator; /// \brief Retrieve an iterator pointing to the first parameter decl. Index: include/clang/Sema/Sema.h =================================================================== --- include/clang/Sema/Sema.h +++ include/clang/Sema/Sema.h @@ -2981,6 +2981,10 @@ void ActOnCapturedRegionStart(SourceLocation Loc, Scope *CurScope, CapturedRegionKind Kind, unsigned NumParams); + typedef std::pair ParamNameType; + void ActOnCapturedRegionStart(SourceLocation Loc, Scope *CurScope, + CapturedRegionKind Kind, + ArrayRef Params); StmtResult ActOnCapturedRegionEnd(Stmt *S); void ActOnCapturedRegionError(); RecordDecl *CreateCapturedStmtRecordDecl(CapturedDecl *&CD, @@ -7100,6 +7104,9 @@ SourceLocation Loc, ArrayRef VarList); + // brief Initialization of captured region for OpenMP parallel region. + void ActOnOpenMPRegionStart(OpenMPDirectiveKind DKind, SourceLocation Loc, + Scope *CurScope); StmtResult ActOnOpenMPExecutableDirective(OpenMPDirectiveKind Kind, ArrayRef Clauses, Stmt *AStmt, Index: lib/CodeGen/CGException.cpp =================================================================== --- lib/CodeGen/CGException.cpp +++ lib/CodeGen/CGException.cpp @@ -490,8 +490,14 @@ return; const FunctionDecl* FD = dyn_cast_or_null(D); - if (FD == 0) + if (FD == 0) { + // Check if CapturedDecl is nothrow and create terminate scope for it. + if (const CapturedDecl* CD = dyn_cast_or_null(D)) { + if (CD->isNothrow()) + EHStack.pushTerminate(); + } return; + } const FunctionProtoType *Proto = FD->getType()->getAs(); if (Proto == 0) return; @@ -558,8 +564,14 @@ return; const FunctionDecl* FD = dyn_cast_or_null(D); - if (FD == 0) + if (FD == 0) { + // Check if CapturedDecl is nothrow and pop terminate scope for it. + if (const CapturedDecl* CD = dyn_cast_or_null(D)) { + if (CD->isNothrow()) + EHStack.popTerminate(); + } return; + } const FunctionProtoType *Proto = FD->getType()->getAs(); if (Proto == 0) return; Index: lib/CodeGen/CGOpenMPRuntime.h =================================================================== --- lib/CodeGen/CGOpenMPRuntime.h +++ lib/CodeGen/CGOpenMPRuntime.h @@ -0,0 +1,139 @@ +//===----- CGOpenMPRuntime.h - Interface to OpenMP Runtimes -----*- C++ -*-===// +// +// The LLVM Compiler Infrastructure +// +// This file is distributed under the University of Illinois Open Source +// License. See LICENSE.TXT for details. +// +//===----------------------------------------------------------------------===// +// +// This provides a class for OpenMP runtime code generation. +// +//===----------------------------------------------------------------------===// + +#ifndef CLANG_CODEGEN_OPENMPRUNTIME_H +#define CLANG_CODEGEN_OPENMPRUNTIME_H + +#include "clang/AST/Type.h" +#include "llvm/ADT/DenseMap.h" +#include "llvm/IR/Type.h" +#include "llvm/IR/Value.h" + +namespace llvm { +class AllocaInst; +class CallInst; +class GlobalVariable; +class Constant; +class Function; +class Module; +class StructLayout; +class FunctionType; +class StructType; +class Type; +class Value; +} + +namespace clang { + +namespace CodeGen { + +class CodeGenFunction; +class CodeGenModule; + +class CGOpenMPRuntime { +public: + /// \brief Values for bit flags used in the ident_t to describe the fields. + /// All enumeric elements are named and described in accordance with the code + /// from http://llvm.org/svn/llvm-project/openmp/trunk/runtime/src/kmp.h + enum OpenMPLocationFlags { + /// \brief Use trampoline for internal microtask. + OMP_IDENT_IMD = 0x01, + /// \brief Use c-style ident structure. + OMP_IDENT_KMPC = 0x02, + /// \brief Atomic reduction option for kmpc_reduce. + OMP_ATOMIC_REDUCE = 0x10, + /// \brief Explicit 'barrier' directive. + OMP_IDENT_BARRIER_EXPL = 0x20, + /// \brief Implicit barrier in code. + OMP_IDENT_BARRIER_IMPL = 0x40, + /// \brief Implicit barrier in 'for' directive. + OMP_IDENT_BARRIER_IMPL_FOR = 0x40, + /// \brief Implicit barrier in 'sections' directive. + OMP_IDENT_BARRIER_IMPL_SECTIONS = 0xC0, + /// \brief Implicit barrier in 'single' directive. + OMP_IDENT_BARRIER_IMPL_SINGLE = 0x140 + }; + enum OpenMPRTLFunction { + // Call to void __kmpc_fork_call(ident_t *loc, kmp_int32 argc, kmpc_micro + // microtask, ...); + OMPRTL__kmpc_fork_call, + // Call to kmp_int32 kmpc_global_thread_num(ident_t *loc); + OMPRTL__kmpc_global_thread_num + }; + +private: + // \brief Describes ident structure that describes a source location. + /// All descriptions are taken from + /// http://llvm.org/svn/llvm-project/openmp/trunk/runtime/src/kmp.h + enum Ident_TFields { + /// \brief might be used in Fortran + Ident_T_Reserved_1, + /// \brief OMP_IDENT_xxx flags; OMP_IDENT_KMPC identifies this union member. + Ident_T_Flags, + /// \brief Not really used in Fortran any more + Ident_T_Reserved_2, + /// \brief Source[4] in Fortran, do not use for C++ + Ident_T_Reserved_3, + /// \brief String describing the source location. The string is composed of + /// semi-colon separated fields which describe the source file, the function + /// and a pair of line numbers that delimit the construct. + Ident_T_PSource + }; + CodeGenModule &CGM; + /// \brief Default const ident_t object used for initialization of all other + /// ident_t objects. + llvm::Constant *DefaultOpenMPPSource; + llvm::StructType *Ident_TTy; + llvm::FunctionType *Kmpc_MicroTy; + llvm::GlobalVariable *DefaultOpenMPLocation; + /// \brief Map of local debug location and functions. + typedef llvm::DenseMap OpenMPLocMapTy; + OpenMPLocMapTy OpenMPLocMap; + /// \brief Map of local gtid and functions. + typedef llvm::DenseMap OpenMPGtidMapTy; + OpenMPGtidMapTy OpenMPGtidMap; + +public: + CGOpenMPRuntime(CodeGenModule &CGM); + ~CGOpenMPRuntime() {} + + /// \brief Generates object of ident_t type with info for source location. + /// \param CGF Reference to current CodeGenFunction. + /// \param Loc Clang source location. + /// \param Flags Flags for OpenMP location. + /// + llvm::Value *GetOpenMPLocation(CodeGenFunction &CGF, SourceLocation Loc, + OpenMPLocationFlags Flags = OMP_IDENT_KMPC); + + /// \brief Generates global thread number value. + /// \param CGF Reference to current CodeGenFunction. + /// \param Loc Clang source location. + /// + llvm::Value *GetOpenMPGlobalThreadNum(CodeGenFunction &CGF, + SourceLocation Loc); + + /// \brief Returns pointer to ident_t type; + llvm::Type *getIdent_TPointerTy(); + + /// \brief Returns pointer to kmpc_micro type; + llvm::Type *getKmpc_MicroPointerTy(); + + /// \brief Returns specified OpenMP runtime function. + /// \param Function OpenMP runtime function. + /// \return Specified function. + llvm::Constant *CreateRuntimeFunction(OpenMPRTLFunction Function); +}; +} +} + +#endif Index: lib/CodeGen/CGOpenMPRuntime.cpp =================================================================== --- lib/CodeGen/CGOpenMPRuntime.cpp +++ lib/CodeGen/CGOpenMPRuntime.cpp @@ -0,0 +1,171 @@ +//===----- CGOpenMPRuntime.cpp - Interface to OpenMP Runtimes -------------===// +// +// The LLVM Compiler Infrastructure +// +// This file is distributed under the University of Illinois Open Source +// License. See LICENSE.TXT for details. +// +//===----------------------------------------------------------------------===// +// +// This provides a class for OpenMP runtime code generation. +// +//===----------------------------------------------------------------------===// + +#include "CGOpenMPRuntime.h" +#include "CodeGenFunction.h" +#include "clang/AST/Decl.h" +#include "llvm/ADT/ArrayRef.h" +#include "llvm/IR/DerivedTypes.h" +#include "llvm/IR/GlobalValue.h" +#include "llvm/IR/Value.h" +#include "llvm/Support/raw_ostream.h" +#include + +using namespace clang; +using namespace CodeGen; + +CGOpenMPRuntime::CGOpenMPRuntime(CodeGenModule &CGM) : CGM(CGM) { + // Initialize default location for psource field of ident_t structure of all + // ident_t objects. Format is ";file;function;line;column;;". Taken from + // http://llvm.org/svn/llvm-project/openmp/trunk/runtime/src/kmp_str.c + DefaultOpenMPPSource = CGM.GetAddrOfConstantCString(";unknown;unknown;0;0;;"); + DefaultOpenMPPSource = + llvm::ConstantExpr::getBitCast(DefaultOpenMPPSource, CGM.Int8PtrTy); + Ident_TTy = llvm::StructType::create( + "ident_t", CGM.Int32Ty /* reserved_1 */, CGM.Int32Ty /* flags */, + CGM.Int32Ty /* reserved_2 */, CGM.Int32Ty /* reserved_3 */, + CGM.Int8PtrTy /* psource */, NULL); + // Build void (*kmpc_micro)(kmp_int32 *global_tid, kmp_int32 *bound_tid,...) + llvm::Type *MicroParams[] = { llvm::PointerType::getUnqual(CGM.Int32Ty), + llvm::PointerType::getUnqual(CGM.Int32Ty) }; + Kmpc_MicroTy = llvm::FunctionType::get(CGM.VoidTy, MicroParams, true); + llvm::Constant *Zero = llvm::ConstantInt::get(CGM.Int32Ty, 0, true); + llvm::Constant *Values[] = { Zero, llvm::ConstantInt::get(CGM.Int32Ty, + OMP_IDENT_KMPC), + Zero, Zero, DefaultOpenMPPSource }; + llvm::Constant *Init = llvm::ConstantStruct::get(Ident_TTy, Values); + DefaultOpenMPLocation = new llvm::GlobalVariable( + CGM.getModule(), Ident_TTy, true, llvm::GlobalValue::PrivateLinkage, Init, + ".kmpc_default_loc.addr"); + DefaultOpenMPLocation->setUnnamedAddr(true); +} + +llvm::Value *CGOpenMPRuntime::GetOpenMPLocation(CodeGenFunction &CGF, + SourceLocation Loc, + OpenMPLocationFlags Flags) { + // If no debug info is generated - return global default location. + if (CGM.getCodeGenOpts().getDebugInfo() == CodeGenOptions::NoDebugInfo) + return DefaultOpenMPLocation; + + assert(CGF.CurFn && "No function in current CodeGenFunction."); + + SmallString<32> Buffer; + llvm::raw_svector_ostream OS(Buffer); + OS << ".kmpc_loc_" << Flags << ".addr"; + + llvm::AllocaInst *AI = 0; + OpenMPLocMapTy::iterator I = OpenMPLocMap.find(CGF.CurFn); + if (I != OpenMPLocMap.end()) { + AI = I->second; + } else { + // Generate "ident_t .kmpc_loc_.addr;" + AI = CGF.CreateTempAlloca(Ident_TTy, OS.str()); + AI->setAlignment(CGM.getDataLayout().getPrefTypeAlignment(Ident_TTy)); + OpenMPLocMap[CGF.CurFn] = AI; + + CGBuilderTy::InsertPointGuard IPG(CGF.Builder); + CGF.Builder.SetInsertPoint(CGF.AllocaInsertPt); + CGF.Builder.CreateMemCpy(AI, DefaultOpenMPLocation, + llvm::ConstantExpr::getSizeOf(Ident_TTy), + CGM.PointerAlignInBytes); + // int32 *flags = &(.kmpc_loc_.addr).flags; + llvm::Value *FlagsField = + CGF.Builder.CreateConstInBoundsGEP2_32(AI, 0, Ident_T_Flags); + // *flags = ; + CGF.Builder.CreateStore(CGF.Builder.getInt32(Flags), FlagsField); + } + + // char **psource = &.kmpc_loc_.addr.psource; + llvm::Value *PSource = + CGF.Builder.CreateConstInBoundsGEP2_32(AI, 0, Ident_T_PSource); + if (CGM.getCodeGenOpts().getDebugInfo() != CodeGenOptions::NoDebugInfo) { + if (Loc.isValid()) { + SmallString<128> Buffer2; + llvm::raw_svector_ostream OS2(Buffer2); + // Build debug location + PresumedLoc PLoc = + CGF.getContext().getSourceManager().getPresumedLoc(Loc); + OS2 << ";" << PLoc.getFilename() << ";"; + if (const FunctionDecl *FD = + dyn_cast_or_null(CGF.CurFuncDecl)) { + OS2 << FD->getQualifiedNameAsString(); + } + OS2 << ";" << PLoc.getLine() << ";" << PLoc.getColumn() << ";;"; + // *psource = ";;;;;;"; + CGF.Builder.CreateStore(CGF.Builder.CreateGlobalStringPtr(OS2.str()), + PSource); + } else { + // *psource = ";unknown;unknown;0;0;;"; + CGF.Builder.CreateStore(DefaultOpenMPPSource, PSource); + } + } + return AI; +} + +llvm::Value *CGOpenMPRuntime::GetOpenMPGlobalThreadNum(CodeGenFunction &CGF, + SourceLocation Loc) { + assert(CGF.CurFn && "No function in current CodeGenFunction."); + + llvm::AllocaInst *AI = 0; + OpenMPGtidMapTy::iterator I = OpenMPGtidMap.find(CGF.CurFn); + if (I != OpenMPGtidMap.end()) { + AI = I->second; + } else { + // Generate "int32 .kmpc_global_thread_num.addr;" + AI = CGF.CreateTempAlloca(CGF.Int32Ty, ".kmpc_global_thread_num.addr"); + AI->setAlignment(CGM.getDataLayout().getPrefTypeAlignment(CGF.Int32Ty)); + + CGBuilderTy::InsertPointGuard IPG(CGF.Builder); + CGF.Builder.SetInsertPoint(CGF.AllocaInsertPt); + llvm::Value *Args[] = { GetOpenMPLocation(CGF, Loc) }; + CGF.Builder.CreateStore( + CGF.EmitRuntimeCall( + CreateRuntimeFunction(OMPRTL__kmpc_global_thread_num), Args), + AI); + } + return CGF.Builder.CreateLoad(AI, ".kmpc_global_thread_num"); +} + +llvm::Type *CGOpenMPRuntime::getIdent_TPointerTy() { + return llvm::PointerType::getUnqual(Ident_TTy); +} + +llvm::Type *CGOpenMPRuntime::getKmpc_MicroPointerTy() { + return llvm::PointerType::getUnqual(Kmpc_MicroTy); +} + +llvm::Constant * +CGOpenMPRuntime::CreateRuntimeFunction(OpenMPRTLFunction Function) { + llvm::Constant *RTLFn = 0; + switch (Function) { + case OMPRTL__kmpc_fork_call: { + // Build void __kmpc_fork_call(ident_t *loc, kmp_int32 argc, kmpc_micro + // microtask, ...); + llvm::Type *TypeParams[] = { getIdent_TPointerTy(), CGM.Int32Ty, + getKmpc_MicroPointerTy() }; + llvm::FunctionType *FnTy = + llvm::FunctionType::get(CGM.VoidTy, TypeParams, true); + RTLFn = CGM.CreateRuntimeFunction(FnTy, "__kmpc_fork_call"); + break; + } + case OMPRTL__kmpc_global_thread_num: { + // Build kmp_int32 __kmpc_global_thread_num(ident_t *loc); + llvm::Type *TypeParams[] = { getIdent_TPointerTy() }; + llvm::FunctionType *FnTy = + llvm::FunctionType::get(CGM.Int32Ty, TypeParams, false); + RTLFn = CGM.CreateRuntimeFunction(FnTy, "__kmpc_global_thread_num"); + break; + } + } + return RTLFn; +} Index: lib/CodeGen/CGStmt.cpp =================================================================== --- lib/CodeGen/CGStmt.cpp +++ lib/CodeGen/CGStmt.cpp @@ -76,7 +76,6 @@ case Stmt::SEHExceptStmtClass: case Stmt::SEHFinallyStmtClass: case Stmt::MSDependentExistsStmtClass: - case Stmt::OMPParallelDirectiveClass: llvm_unreachable("invalid statement class to emit generically"); case Stmt::NullStmtClass: case Stmt::CompoundStmtClass: @@ -173,6 +172,9 @@ case Stmt::SEHTryStmtClass: EmitSEHTryStmt(cast(*S)); break; + case Stmt::OMPParallelDirectiveClass: + EmitOMPParallelDirective(cast(*S)); + break; } } @@ -1919,6 +1921,12 @@ return F; } +llvm::Value * +CodeGenFunction::GenerateCapturedStmtArgument(const CapturedStmt &S) { + LValue CapStruct = InitCapturedStruct(*this, S); + return CapStruct.getAddress(); +} + /// Creates the outlined function for a CapturedStmt. llvm::Function * CodeGenFunction::GenerateCapturedStmtFunction(const CapturedDecl *CD, Index: lib/CodeGen/CGStmtOpenMP.cpp =================================================================== --- lib/CodeGen/CGStmtOpenMP.cpp +++ lib/CodeGen/CGStmtOpenMP.cpp @@ -0,0 +1,50 @@ +//===--- CGStmtOpenMP.cpp - Emit LLVM Code from Statements ----------------===// +// +// The LLVM Compiler Infrastructure +// +// This file is distributed under the University of Illinois Open Source +// License. See LICENSE.TXT for details. +// +//===----------------------------------------------------------------------===// +// +// This contains code to emit OpenMP nodes as LLVM code. +// +//===----------------------------------------------------------------------===// + +#include "CGOpenMPRuntime.h" +#include "CodeGenFunction.h" +#include "CodeGenModule.h" +#include "clang/AST/Stmt.h" +#include "clang/AST/StmtOpenMP.h" +using namespace clang; +using namespace CodeGen; + +//===----------------------------------------------------------------------===// +// OpenMP Directive Emission +//===----------------------------------------------------------------------===// + +void CodeGenFunction::EmitOMPParallelDirective(const OMPParallelDirective &S) { + const CapturedStmt *CS = cast(S.getAssociatedStmt()); + llvm::Value *CapturedStruct = GenerateCapturedStmtArgument(*CS); + + CodeGenFunction CGF(CGM, true); + llvm::Value *OutlinedFn; + { + CGCapturedStmtInfo CGInfo(*CS, CS->getCapturedRegionKind()); + CGF.CapturedStmtInfo = &CGInfo; + OutlinedFn = CGF.GenerateCapturedStmtFunction( + CS->getCapturedDecl(), CS->getCapturedRecordDecl(), CS->getLocStart()); + } + + // Build call __kmpc_fork_call(loc, 1, microtask, captured_struct/*context*/) + llvm::Value *Args[] = { + CGM.getOpenMPRuntime().GetOpenMPLocation(*this, S.getLocStart()), + Builder.getInt32(1), + Builder.CreateBitCast(OutlinedFn, + CGM.getOpenMPRuntime().getKmpc_MicroPointerTy()), + EmitCastToVoidPtr(CapturedStruct) + }; + llvm::Constant *RTLFn = CGM.getOpenMPRuntime().CreateRuntimeFunction( + CGOpenMPRuntime::OMPRTL__kmpc_fork_call); + EmitRuntimeCall(RTLFn, Args); +} Index: lib/CodeGen/CMakeLists.txt =================================================================== --- lib/CodeGen/CMakeLists.txt +++ lib/CodeGen/CMakeLists.txt @@ -42,9 +42,11 @@ CGObjCMac.cpp CGObjCRuntime.cpp CGOpenCLRuntime.cpp + CGOpenMPRuntime.cpp CGRTTI.cpp CGRecordLayoutBuilder.cpp CGStmt.cpp + CGStmtOpenMP.cpp CGVTT.cpp CGVTables.cpp CodeGenABITypes.cpp Index: lib/CodeGen/CodeGenFunction.h =================================================================== --- lib/CodeGen/CodeGenFunction.h +++ lib/CodeGen/CodeGenFunction.h @@ -1872,6 +1872,9 @@ llvm::Function *GenerateCapturedStmtFunction(const CapturedDecl *CD, const RecordDecl *RD, SourceLocation Loc); + llvm::Value *GenerateCapturedStmtArgument(const CapturedStmt &S); + + void EmitOMPParallelDirective(const OMPParallelDirective &S); //===--------------------------------------------------------------------===// // LValue Expression Emission Index: lib/CodeGen/CodeGenModule.h =================================================================== --- lib/CodeGen/CodeGenModule.h +++ lib/CodeGen/CodeGenModule.h @@ -82,6 +82,7 @@ class CGDebugInfo; class CGObjCRuntime; class CGOpenCLRuntime; + class CGOpenMPRuntime; class CGCUDARuntime; class BlockFieldFlags; class FunctionArgList; @@ -254,6 +255,7 @@ CGObjCRuntime* ObjCRuntime; CGOpenCLRuntime* OpenCLRuntime; + CGOpenMPRuntime* OpenMPRuntime; CGCUDARuntime* CUDARuntime; CGDebugInfo* DebugInfo; ARCEntrypoints *ARCData; @@ -402,6 +404,7 @@ void createObjCRuntime(); void createOpenCLRuntime(); + void createOpenMPRuntime(); void createCUDARuntime(); bool isTriviallyRecursive(const FunctionDecl *F); @@ -465,6 +468,12 @@ return *OpenCLRuntime; } + /// getOpenMPRuntime() - Return a reference to the configured OpenMP runtime. + CGOpenMPRuntime &getOpenMPRuntime() { + assert(OpenMPRuntime != 0); + return *OpenMPRuntime; + } + /// getCUDARuntime() - Return a reference to the configured CUDA runtime. CGCUDARuntime &getCUDARuntime() { assert(CUDARuntime != 0); Index: lib/CodeGen/CodeGenModule.cpp =================================================================== --- lib/CodeGen/CodeGenModule.cpp +++ lib/CodeGen/CodeGenModule.cpp @@ -18,6 +18,7 @@ #include "CGDebugInfo.h" #include "CGObjCRuntime.h" #include "CGOpenCLRuntime.h" +#include "CGOpenMPRuntime.h" #include "CodeGenFunction.h" #include "CodeGenPGO.h" #include "CodeGenTBAA.h" @@ -76,8 +77,8 @@ Diags(diags), TheDataLayout(TD), Target(C.getTargetInfo()), ABI(createCXXABI(*this)), VMContext(M.getContext()), TBAA(0), TheTargetCodeGenInfo(0), Types(*this), VTables(*this), ObjCRuntime(0), - OpenCLRuntime(0), CUDARuntime(0), DebugInfo(0), ARCData(0), - NoObjCARCExceptionsMetadata(0), RRData(0), PGOData(0), + OpenCLRuntime(0), OpenMPRuntime(0), CUDARuntime(0), DebugInfo(0), + ARCData(0), NoObjCARCExceptionsMetadata(0), RRData(0), PGOData(0), CFConstantStringClassRef(0), ConstantStringClassRef(0), NSConstantStringType(0), NSConcreteGlobalBlock(0), NSConcreteStackBlock(0), BlockObjectAssign(0), @@ -111,6 +112,8 @@ createObjCRuntime(); if (LangOpts.OpenCL) createOpenCLRuntime(); + if (LangOpts.OpenMP) + createOpenMPRuntime(); if (LangOpts.CUDA) createCUDARuntime(); @@ -140,6 +143,7 @@ CodeGenModule::~CodeGenModule() { delete ObjCRuntime; delete OpenCLRuntime; + delete OpenMPRuntime; delete CUDARuntime; delete TheTargetCodeGenInfo; delete TBAA; @@ -171,6 +175,10 @@ OpenCLRuntime = new CGOpenCLRuntime(*this); } +void CodeGenModule::createOpenMPRuntime() { + OpenMPRuntime = new CGOpenMPRuntime(*this); +} + void CodeGenModule::createCUDARuntime() { CUDARuntime = CreateNVCUDARuntime(*this); } Index: lib/Parse/ParseOpenMP.cpp =================================================================== --- lib/Parse/ParseOpenMP.cpp +++ lib/Parse/ParseOpenMP.cpp @@ -11,8 +11,9 @@ /// //===----------------------------------------------------------------------===// -#include "clang/AST/ASTConsumer.h" #include "RAIIObjectsForParser.h" +#include "clang/AST/ASTConsumer.h" +#include "clang/AST/ASTContext.h" #include "clang/AST/StmtOpenMP.h" #include "clang/Parse/ParseDiagnostic.h" #include "clang/Parse/Parser.h" @@ -146,7 +147,7 @@ { // The body is a block scope like in Lambdas and Blocks. Sema::CompoundScopeRAII CompoundScope(Actions); - Actions.ActOnCapturedRegionStart(Loc, getCurScope(), CR_OpenMP, 1); + Actions.ActOnOpenMPRegionStart(DKind, Loc, getCurScope()); Actions.ActOnStartOfCompoundStmt(); // Parse statement AssociatedStmt = ParseStatement(); Index: lib/Sema/SemaOpenMP.cpp =================================================================== --- lib/Sema/SemaOpenMP.cpp +++ lib/Sema/SemaOpenMP.cpp @@ -1,4 +1,4 @@ -//===--- SemaOpenMP.cpp - Semantic Analysis for OpenMP constructs ----------===// +//===--- SemaOpenMP.cpp - Semantic Analysis for OpenMP constructs ---------===// // // The LLVM Compiler Infrastructure // @@ -12,13 +12,14 @@ /// //===----------------------------------------------------------------------===// -#include "clang/Basic/OpenMPKinds.h" +#include "clang/AST/ASTContext.h" #include "clang/AST/Decl.h" #include "clang/AST/DeclCXX.h" #include "clang/AST/DeclOpenMP.h" #include "clang/AST/StmtCXX.h" #include "clang/AST/StmtOpenMP.h" #include "clang/AST/StmtVisitor.h" +#include "clang/Basic/OpenMPKinds.h" #include "clang/Lex/Preprocessor.h" #include "clang/Sema/Initialization.h" #include "clang/Sema/Lookup.h" @@ -663,6 +664,29 @@ }; } +void Sema::ActOnOpenMPRegionStart(OpenMPDirectiveKind DKind, SourceLocation Loc, + Scope *CurScope) { + switch (DKind) { + case OMPD_parallel: { + QualType KmpInt32Ty = Context.getIntTypeForBitwidth(32, 1); + QualType KmpInt32PtrTy = Context.getPointerType(KmpInt32Ty); + Sema::ParamNameType Params[3] = { + std::make_pair(".global_tid.", KmpInt32PtrTy), + std::make_pair(".bound_tid.", KmpInt32PtrTy), + std::make_pair(StringRef(), QualType()) // __context with shared vars + }; + ActOnCapturedRegionStart(Loc, CurScope, CR_OpenMP, Params); + break; + } + case OMPD_threadprivate: + case OMPD_task: + llvm_unreachable("OpenMP Directive is not allowed"); + case OMPD_unknown: + case NUM_OPENMP_DIRECTIVES: + llvm_unreachable("Unknown OpenMP directive"); + } +} + StmtResult Sema::ActOnOpenMPExecutableDirective(OpenMPDirectiveKind Kind, ArrayRef Clauses, Stmt *AStmt, @@ -715,6 +739,11 @@ Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc) { + assert(AStmt && isa(AStmt) && "Captured statement expected"); + CapturedStmt *CS = cast(AStmt); + // Mark captured decl as nothrow + CS->getCapturedDecl()->setNothrow(); + getCurFunction()->setHasBranchProtectedScope(); return Owned(OMPParallelDirective::Create(Context, StartLoc, EndLoc, Index: lib/Sema/SemaStmt.cpp =================================================================== --- lib/Sema/SemaStmt.cpp +++ lib/Sema/SemaStmt.cpp @@ -3284,20 +3284,9 @@ RD->setImplicit(); RD->startDefinition(); + assert(NumParams > 0 && "CapturedStmt requires context parameter"); CD = CapturedDecl::Create(Context, CurContext, NumParams); DC->addDecl(CD); - - // Build the context parameter - assert(NumParams > 0 && "CapturedStmt requires context parameter"); - DC = CapturedDecl::castToDeclContext(CD); - IdentifierInfo *VarName = &Context.Idents.get("__context"); - QualType ParamType = Context.getPointerType(Context.getTagDeclType(RD)); - ImplicitParamDecl *Param - = ImplicitParamDecl::Create(Context, DC, Loc, VarName, ParamType); - DC->addDecl(Param); - - CD->setContextParam(Param); - return RD; } @@ -3332,6 +3321,58 @@ CapturedDecl *CD = 0; RecordDecl *RD = CreateCapturedStmtRecordDecl(CD, Loc, NumParams); + // Build the context parameter + DeclContext *DC = CapturedDecl::castToDeclContext(CD); + IdentifierInfo *ParamName = &Context.Idents.get("__context"); + QualType ParamType = Context.getPointerType(Context.getTagDeclType(RD)); + ImplicitParamDecl *Param + = ImplicitParamDecl::Create(Context, DC, Loc, ParamName, ParamType); + DC->addDecl(Param); + + CD->setContextParam(0, Param); + + // Enter the capturing scope for this captured region. + PushCapturedRegionScope(CurScope, CD, RD, Kind); + + if (CurScope) + PushDeclContext(CurScope, CD); + else + CurContext = CD; + + PushExpressionEvaluationContext(PotentiallyEvaluated); +} + +void Sema::ActOnCapturedRegionStart(SourceLocation Loc, Scope *CurScope, + CapturedRegionKind Kind, + ArrayRef Params) { + CapturedDecl *CD = 0; + RecordDecl *RD = CreateCapturedStmtRecordDecl(CD, Loc, Params.size()); + + // Build the context parameter + DeclContext *DC = CapturedDecl::castToDeclContext(CD); + bool ContextIsFound = false; + unsigned ParamNum = 0; + for (ArrayRef::iterator I = Params.begin(), E = Params.end(); + I != E; ++I, ++ParamNum) { + if (I->second.isNull()) { + assert(!ContextIsFound && + "null type has been found already for '__context' parameter"); + IdentifierInfo *ParamName = &Context.Idents.get("__context"); + QualType ParamType = Context.getPointerType(Context.getTagDeclType(RD)); + ImplicitParamDecl *Param + = ImplicitParamDecl::Create(Context, DC, Loc, ParamName, ParamType); + DC->addDecl(Param); + CD->setContextParam(ParamNum, Param); + ContextIsFound = true; + } else { + IdentifierInfo *ParamName = &Context.Idents.get(I->first); + ImplicitParamDecl *Param + = ImplicitParamDecl::Create(Context, DC, Loc, ParamName, I->second); + DC->addDecl(Param); + CD->setParam(ParamNum, Param); + } + } + assert(ContextIsFound && "no null type for '__context' parameter"); // Enter the capturing scope for this captured region. PushCapturedRegionScope(CurScope, CD, RD, Kind); Index: lib/Sema/TreeTransform.h =================================================================== --- lib/Sema/TreeTransform.h +++ lib/Sema/TreeTransform.h @@ -9777,9 +9777,22 @@ StmtResult TreeTransform::TransformCapturedStmt(CapturedStmt *S) { SourceLocation Loc = S->getLocStart(); - unsigned NumParams = S->getCapturedDecl()->getNumParams(); + CapturedDecl *CD = S->getCapturedDecl(); + unsigned NumParams = CD->getNumParams(); + unsigned ContextParamPos = CD->getContextParamPosition(); + SmallVector Params; + for (unsigned I = 0; I < NumParams; ++I) { + if (I != ContextParamPos) { + Params.push_back( + std::make_pair( + CD->getParam(I)->getName(), + getDerived().TransformType(CD->getParam(I)->getType()))); + } else { + Params.push_back(std::make_pair(StringRef(), QualType())); + } + } getSema().ActOnCapturedRegionStart(Loc, /*CurScope*/0, - S->getCapturedRegionKind(), NumParams); + S->getCapturedRegionKind(), Params); StmtResult Body = getDerived().TransformStmt(S->getCapturedStmt()); if (Body.isInvalid()) { Index: lib/Serialization/ASTReaderDecl.cpp =================================================================== --- lib/Serialization/ASTReaderDecl.cpp +++ lib/Serialization/ASTReaderDecl.cpp @@ -1077,9 +1077,15 @@ void ASTDeclReader::VisitCapturedDecl(CapturedDecl *CD) { VisitDecl(CD); + unsigned ContextParamPos = Record[Idx++]; + CD->setNothrow(Record[Idx++] != 0); // Body is set by VisitCapturedStmt. - for (unsigned i = 0; i < CD->NumParams; ++i) - CD->setParam(i, ReadDeclAs(Record, Idx)); + for (unsigned I = 0; I < CD->NumParams; ++I) { + if (I != ContextParamPos) + CD->setParam(I, ReadDeclAs(Record, Idx)); + else + CD->setContextParam(I, ReadDeclAs(Record, Idx)); + } } void ASTDeclReader::VisitLinkageSpecDecl(LinkageSpecDecl *D) { Index: lib/Serialization/ASTWriterDecl.cpp =================================================================== --- lib/Serialization/ASTWriterDecl.cpp +++ lib/Serialization/ASTWriterDecl.cpp @@ -853,9 +853,11 @@ void ASTDeclWriter::VisitCapturedDecl(CapturedDecl *CD) { Record.push_back(CD->getNumParams()); VisitDecl(CD); + Record.push_back(CD->getContextParamPosition()); + Record.push_back(CD->isNothrow() ? 1 : 0); // Body is stored by VisitCapturedStmt. - for (unsigned i = 0; i < CD->getNumParams(); ++i) - Writer.AddDeclRef(CD->getParam(i), Record); + for (unsigned I = 0; I < CD->getNumParams(); ++I) + Writer.AddDeclRef(CD->getParam(I), Record); Code = serialization::DECL_CAPTURED; } Index: test/OpenMP/parallel_codegen.cpp =================================================================== --- test/OpenMP/parallel_codegen.cpp +++ test/OpenMP/parallel_codegen.cpp @@ -0,0 +1,150 @@ +// RUN: %clang_cc1 -verify -fopenmp -x c++ -emit-llvm %s -fexceptions -fcxx-exceptions -o - | FileCheck %s +// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple x86_64-unknown-unknown -fexceptions -fcxx-exceptions -emit-pch -o %t %s +// RUN: %clang_cc1 -fopenmp -x c++ -triple x86_64-unknown-unknown -fexceptions -fcxx-exceptions -g -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix=CHECK-DEBUG %s +// expected-no-diagnostics + +#ifndef HEADER +#define HEADER + +// CHECK-DAG: %ident_t = type { i32, i32, i32, i32, i8* } +// CHECK-DAG: %struct.anon = type { i32* } +// CHECK-DAG: %struct.anon.0 = type { i8*** } +// CHECK-DAG: @.str = private unnamed_addr constant [23 x i8] c";unknown;unknown;0;0;;\00" +// CHECK-DAG: @.kmpc_default_loc.addr = private unnamed_addr constant %ident_t { i32 0, i32 2, i32 0, i32 0, i8* getelementptr inbounds ([23 x i8]* @.str, i32 0, i32 0) } +// CHECK-DEBUG-DAG: %ident_t = type { i32, i32, i32, i32, i8* } +// CHECK-DEBUG-DAG: %struct.anon = type { i32* } +// CHECK-DEBUG-DAG: %struct.anon.0 = type { i8*** } +// CHECK-DEBUG-DAG: @.str = private unnamed_addr constant [23 x i8] c";unknown;unknown;0;0;;\00" +// CHECK-DEBUG-DAG: @.kmpc_default_loc.addr = private unnamed_addr constant %ident_t { i32 0, i32 2, i32 0, i32 0, i8* getelementptr inbounds ([23 x i8]* @.str, i32 0, i32 0) } +// CHECK-DEBUG-DAG: [[LOC1:@.+]] = private unnamed_addr constant [{{.+}} x i8] c";{{.*}}parallel_codegen.cpp;main;[[@LINE+14]];9;;\00" +// CHECK-DEBUG-DAG: [[LOC2:@.+]] = private unnamed_addr constant [{{.+}} x i8] c";{{.*}}parallel_codegen.cpp;tmain;[[@LINE+7]];9;;\00" + +template +void foo(T argc) {} + +template +int tmain(T argc) { +#pragma omp parallel + foo(argc); + return 0; +} + +int main (int argc, char **argv) { +#pragma omp parallel + foo(argc); + return tmain(argv); +} + +// CHECK-LABEL: define i32 @main(i32 %argc, i8** %argv) +// CHECK: %agg.captured = alloca %struct.anon +// CHECK: [[ARGC_REF:%.+]] = getelementptr inbounds %struct.anon* %agg.captured, i32 0, i32 0 +// CHECK-NEXT: store i32* %argc.addr, i32** [[ARGC_REF]] +// CHECK-NEXT: [[BITCAST:%.+]] = bitcast %struct.anon* %agg.captured to i8* +// CHECK-NEXT: call void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)* @__kmpc_fork_call(%ident_t* @.kmpc_default_loc.addr, i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, %struct.anon*)* @__captured_stmt to void (i32*, i32*, ...)*), i8* [[BITCAST]]) +// CHECK-NEXT: [[ARGV:%.+]] = load i8*** %argv.addr, align 8 +// CHECK-NEXT: [[RET:%.+]] = call i32 @_Z5tmainIPPcEiT_(i8** [[ARGV]]) +// CHECK-NEXT: ret i32 [[RET]] +// CHECK-NEXT: } +// CHECK-DEBUG-LABEL: define i32 @main(i32 %argc, i8** %argv) +// CHECK-DEBUG-DAG: %agg.captured = alloca %struct.anon +// CHECK-DEBUG-DAG: %.kmpc_loc_2.addr = alloca %ident_t +// CHECK-DEBUG: [[KMPC_LOC_VOIDPTR:%.+]] = bitcast %ident_t* %.kmpc_loc_2.addr to i8* +// CHECK-DEBUG-NEXT: [[KMPC_DEFAULT_LOC_VOIDPTR:%.+]] = bitcast %ident_t* @.kmpc_default_loc.addr to i8* +// CHECK-DEBUG-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* [[KMPC_LOC_VOIDPTR]], i8* [[KMPC_DEFAULT_LOC_VOIDPTR]], i64 ptrtoint (%ident_t* getelementptr (%ident_t* null, i32 1) to i64), i32 8, i1 false) +// CHECK-DEBUG: [[KMPC_LOC_FLAGS_REF:%.+]] = getelementptr inbounds %ident_t* %.kmpc_loc_2.addr, i32 0, i32 1 +// CHECK-DEBUG-NEXT: store i32 2, i32* [[KMPC_LOC_FLAGS_REF]] +// CHECK-DEBUG: [[ARGC_REF:%.+]] = getelementptr inbounds %struct.anon* %agg.captured, i32 0, i32 0 +// CHECK-DEBUG-NEXT: store i32* %argc.addr, i32** [[ARGC_REF]] +// CHECK-DEBUG-NEXT: [[KMPC_LOC_PSOURCE_REF:%.+]] = getelementptr inbounds %ident_t* %.kmpc_loc_2.addr, i32 0, i32 4 +// CHECK-DEBUG-NEXT: store i8* getelementptr inbounds ([{{.+}} x i8]* [[LOC1]], i32 0, i32 0), i8** [[KMPC_LOC_PSOURCE_REF]] +// CHECK-DEBUG-NEXT: [[BITCAST:%.+]] = bitcast %struct.anon* %agg.captured to i8* +// CHECK-DEBUG-NEXT: call void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)* @__kmpc_fork_call(%ident_t* %.kmpc_loc_2.addr, i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, %struct.anon*)* @__captured_stmt to void (i32*, i32*, ...)*), i8* [[BITCAST]]) +// CHECK-DEBUG-NEXT: [[ARGV:%.+]] = load i8*** %argv.addr, align 8 +// CHECK-DEBUG-NEXT: [[RET:%.+]] = call i32 @_Z5tmainIPPcEiT_(i8** [[ARGV]]) +// CHECK-DEBUG-NEXT: ret i32 [[RET]] +// CHECK-DEBUG-NEXT: } + +// CHECK-LABEL: define internal void @__captured_stmt(i32* %.global_tid., i32* %.bound_tid., %struct.anon* %__context) +// CHECK: %__context.addr = alloca %struct.anon* +// CHECK: store %struct.anon* %__context, %struct.anon** %__context.addr +// CHECK-NEXT: [[CONTEXT_PTR:%.+]] = load %struct.anon** %__context.addr +// CHECK-NEXT: [[ARGC_PTR_REF:%.+]] = getelementptr inbounds %struct.anon* [[CONTEXT_PTR]], i32 0, i32 0 +// CHECK-NEXT: [[ARGC_REF:%.+]] = load i32** [[ARGC_PTR_REF]] +// CHECK-NEXT: [[ARGC:%.+]] = load i32* [[ARGC_REF]] +// CHECK-NEXT: invoke void @_Z3fooIiEvT_(i32 [[ARGC]]) +// CHECK: ret void +// CHECK: call void @__clang_call_terminate(i8* +// CHECK-NEXT: unreachable +// CHECK-NEXT: } +// CHECK-DEBUG-LABEL: define internal void @__captured_stmt(i32* %.global_tid., i32* %.bound_tid., %struct.anon* %__context) +// CHECK-DEBUG: %__context.addr = alloca %struct.anon* +// CHECK-DEBUG: store %struct.anon* %__context, %struct.anon** %__context.addr +// CHECK-DEBUG: [[CONTEXT_PTR:%.+]] = load %struct.anon** %__context.addr +// CHECK-DEBUG-NEXT: [[ARGC_PTR_REF:%.+]] = getelementptr inbounds %struct.anon* [[CONTEXT_PTR]], i32 0, i32 0 +// CHECK-DEBUG-NEXT: [[ARGC_REF:%.+]] = load i32** [[ARGC_PTR_REF]] +// CHECK-DEBUG-NEXT: [[ARGC:%.+]] = load i32* [[ARGC_REF]] +// CHECK-DEBUG-NEXT: invoke void @_Z3fooIiEvT_(i32 [[ARGC]]) +// CHECK-DEBUG: ret void +// CHECK-DEBUG: call void @__clang_call_terminate(i8* +// CHECK-DEBUG-NEXT: unreachable +// CHECK-DEBUG-NEXT: } + +// CHECK-DAG: define linkonce_odr void @_Z3fooIiEvT_(i32 %argc) +// CHECK-DAG: declare void @__kmpc_fork_call(%ident_t*, i32, void (i32*, i32*, ...)*, ...) +// CHECK-DEBUG-DAG: define linkonce_odr void @_Z3fooIiEvT_(i32 %argc) +// CHECK-DEBUG-DAG: declare void @__kmpc_fork_call(%ident_t*, i32, void (i32*, i32*, ...)*, ...) + +// CHECK-LABEL: define linkonce_odr i32 @_Z5tmainIPPcEiT_(i8** %argc) +// CHECK: %agg.captured = alloca %struct.anon.0 +// CHECK: [[ARGC_REF:%.+]] = getelementptr inbounds %struct.anon.0* %agg.captured, i32 0, i32 0 +// CHECK-NEXT: store i8*** %argc.addr, i8**** [[ARGC_REF]] +// CHECK-NEXT: [[BITCAST:%.+]] = bitcast %struct.anon.0* %agg.captured to i8* +// CHECK-NEXT: call void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)* @__kmpc_fork_call(%ident_t* @.kmpc_default_loc.addr, i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, %struct.anon.0*)* @__captured_stmt1 to void (i32*, i32*, ...)*), i8* [[BITCAST]]) +// CHECK-NEXT: ret i32 0 +// CHECK-NEXT: } +// CHECK-DEBUG-LABEL: define linkonce_odr i32 @_Z5tmainIPPcEiT_(i8** %argc) +// CHECK-DEBUG-DAG: %agg.captured = alloca %struct.anon.0 +// CHECK-DEBUG-DAG: %.kmpc_loc_2.addr = alloca %ident_t +// CHECK-DEBUG: [[KMPC_LOC_VOIDPTR:%.+]] = bitcast %ident_t* %.kmpc_loc_2.addr to i8* +// CHECK-DEBUG-NEXT: [[KMPC_DEFAULT_LOC_VOIDPTR:%.+]] = bitcast %ident_t* @.kmpc_default_loc.addr to i8* +// CHECK-DEBUG-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* [[KMPC_LOC_VOIDPTR]], i8* [[KMPC_DEFAULT_LOC_VOIDPTR]], i64 ptrtoint (%ident_t* getelementptr (%ident_t* null, i32 1) to i64), i32 8, i1 false) +// CHECK-DEBUG: [[KMPC_LOC_FLAGS_REF:%.+]] = getelementptr inbounds %ident_t* %.kmpc_loc_2.addr, i32 0, i32 1 +// CHECK-DEBUG-NEXT: store i32 2, i32* [[KMPC_LOC_FLAGS_REF]] +// CHECK-DEBUG: [[ARGC_REF:%.+]] = getelementptr inbounds %struct.anon.0* %agg.captured, i32 0, i32 0 +// CHECK-DEBUG-NEXT: store i8*** %argc.addr, i8**** [[ARGC_REF]] +// CHECK-DEBUG-NEXT: [[KMPC_LOC_PSOURCE_REF:%.+]] = getelementptr inbounds %ident_t* %.kmpc_loc_2.addr, i32 0, i32 4 +// CHECK-DEBUG-NEXT: store i8* getelementptr inbounds ([{{.+}} x i8]* [[LOC2]], i32 0, i32 0), i8** [[KMPC_LOC_PSOURCE_REF]] +// CHECK-DEBUG-NEXT: [[BITCAST:%.+]] = bitcast %struct.anon.0* %agg.captured to i8* +// CHECK-DEBUG-NEXT: call void (%ident_t*, i32, void (i32*, i32*, ...)*, ...)* @__kmpc_fork_call(%ident_t* %.kmpc_loc_2.addr, i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, %struct.anon.0*)* @__captured_stmt1 to void (i32*, i32*, ...)*), i8* [[BITCAST]]) +// CHECK-DEBUG-NEXT: ret i32 0 +// CHECK-DEBUG-NEXT: } + +// CHECK-LABEL: define internal void @__captured_stmt1(i32* %.global_tid., i32* %.bound_tid., %struct.anon.0* %__context) +// CHECK: %__context.addr = alloca %struct.anon.0*, align 8 +// CHECK: store %struct.anon.0* %__context, %struct.anon.0** %__context.addr, align 8 +// CHECK-NEXT: [[CONTEXT_PTR:%.+]] = load %struct.anon.0** %__context.addr +// CHECK-NEXT: [[ARGC_PTR_REF:%.+]] = getelementptr inbounds %struct.anon.0* [[CONTEXT_PTR]], i32 0, i32 0 +// CHECK-NEXT: [[ARGC_REF:%.+]] = load i8**** [[ARGC_PTR_REF]] +// CHECK-NEXT: [[ARGC:%.+]] = load i8*** [[ARGC_REF]] +// CHECK-NEXT: invoke void @_Z3fooIPPcEvT_(i8** [[ARGC]]) +// CHECK: ret void +// CHECK: call void @__clang_call_terminate(i8* +// CHECK-NEXT: unreachable +// CHECK-NEXT: } +// CHECK-DEBUG-LABEL: define internal void @__captured_stmt1(i32* %.global_tid., i32* %.bound_tid., %struct.anon.0* %__context) +// CHECK-DEBUG: %__context.addr = alloca %struct.anon.0*, align 8 +// CHECK-DEBUG: store %struct.anon.0* %__context, %struct.anon.0** %__context.addr, align 8 +// CHECK-DEBUG: [[CONTEXT_PTR:%.+]] = load %struct.anon.0** %__context.addr +// CHECK-DEBUG-NEXT: [[ARGC_PTR_REF:%.+]] = getelementptr inbounds %struct.anon.0* [[CONTEXT_PTR]], i32 0, i32 0 +// CHECK-DEBUG-NEXT: [[ARGC_REF:%.+]] = load i8**** [[ARGC_PTR_REF]] +// CHECK-DEBUG-NEXT: [[ARGC:%.+]] = load i8*** [[ARGC_REF]] +// CHECK-DEBUG-NEXT: invoke void @_Z3fooIPPcEvT_(i8** [[ARGC]]) +// CHECK-DEBUG: ret void +// CHECK-DEBUG: call void @__clang_call_terminate(i8* +// CHECK-DEBUG-NEXT: unreachable +// CHECK-DEBUG-NEXT: } + +// CHECK: define linkonce_odr void @_Z3fooIPPcEvT_(i8** %argc) +// CHECK-DEBUG: define linkonce_odr void @_Z3fooIPPcEvT_(i8** %argc) + +#endif