llvmorg-github-actions[bot] wrote:
<!--LLVM PR SUMMARY COMMENT--> @llvm/pr-subscribers-clang-codegen Author: Bill Wendling (bwendling) <details> <summary>Changes</summary> Split up massive function into smaller, easier-to-digest chunks. --- Patch is 45.68 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/196885.diff 2 Files Affected: - (modified) clang/lib/CodeGen/CGStmt.cpp (+362-306) - (modified) clang/lib/CodeGen/CodeGenFunction.h (+84) ``````````diff diff --git a/clang/lib/CodeGen/CGStmt.cpp b/clang/lib/CodeGen/CGStmt.cpp index 7b6035a6968b1..09c2f3e3106f7 100644 --- a/clang/lib/CodeGen/CGStmt.cpp +++ b/clang/lib/CodeGen/CGStmt.cpp @@ -2605,32 +2605,29 @@ static llvm::MDNode *getAsmSrcLocInfo(const StringLiteral *Str, return llvm::MDNode::get(CGF.getLLVMContext(), Locs); } -static void UpdateAsmCallInst(llvm::CallBase &Result, bool HasSideEffect, - bool HasUnwindClobber, bool ReadOnly, - bool ReadNone, bool NoMerge, bool NoConvergent, - const AsmStmt &S, - const std::vector<llvm::Type *> &ResultRegTypes, - const std::vector<llvm::Type *> &ArgElemTypes, - CodeGenFunction &CGF, - std::vector<llvm::Value *> &RegResults) { +void CodeGenFunction::UpdateAsmCallInst( + const AsmStmt &S, llvm::CallBase &Result, const AsmConstraintsInfo &AsmInfo, + bool HasSideEffect, bool HasUnwindClobber, bool NoMerge, bool NoConvergent, + std::vector<llvm::Value *> &RegResults) { if (!HasUnwindClobber) Result.addFnAttr(llvm::Attribute::NoUnwind); if (NoMerge) Result.addFnAttr(llvm::Attribute::NoMerge); + // Attach readnone and readonly attributes. if (!HasSideEffect) { - if (ReadNone) + if (AsmInfo.ReadNone) Result.setDoesNotAccessMemory(); - else if (ReadOnly) + else if (AsmInfo.ReadOnly) Result.setOnlyReadsMemory(); } // Add elementtype attribute for indirect constraints. - for (auto Pair : llvm::enumerate(ArgElemTypes)) { + for (auto Pair : llvm::enumerate(AsmInfo.ArgElemTypes)) { if (Pair.value()) { auto Attr = llvm::Attribute::get( - CGF.getLLVMContext(), llvm::Attribute::ElementType, Pair.value()); + getLLVMContext(), llvm::Attribute::ElementType, Pair.value()); Result.addParamAttr(Pair.index(), Attr); } } @@ -2641,79 +2638,75 @@ static void UpdateAsmCallInst(llvm::CallBase &Result, bool HasSideEffect, if (const auto *gccAsmStmt = dyn_cast<GCCAsmStmt>(&S); gccAsmStmt && (SL = dyn_cast<StringLiteral>(gccAsmStmt->getAsmStringExpr()))) { - Result.setMetadata("srcloc", getAsmSrcLocInfo(SL, CGF)); + Result.setMetadata("srcloc", getAsmSrcLocInfo(SL, *this)); } else { // At least put the line number on MS inline asm blobs and GCC asm constexpr // strings. llvm::Constant *Loc = - llvm::ConstantInt::get(CGF.Int64Ty, S.getAsmLoc().getRawEncoding()); + llvm::ConstantInt::get(Int64Ty, S.getAsmLoc().getRawEncoding()); Result.setMetadata("srcloc", - llvm::MDNode::get(CGF.getLLVMContext(), + llvm::MDNode::get(getLLVMContext(), llvm::ConstantAsMetadata::get(Loc))); } // Make inline-asm calls Key for the debug info feature Key Instructions. - CGF.addInstToNewSourceAtom(&Result, nullptr); + addInstToNewSourceAtom(&Result, nullptr); - if (!NoConvergent && CGF.getLangOpts().assumeFunctionsAreConvergent()) + if (!NoConvergent && getLangOpts().assumeFunctionsAreConvergent()) // Conservatively, mark all inline asm blocks in CUDA or OpenCL as // convergent (meaning, they may call an intrinsically convergent op, such // as bar.sync, and so can't have certain optimizations applied around // them) unless it's explicitly marked 'noconvergent'. Result.addFnAttr(llvm::Attribute::Convergent); // Extract all of the register value results from the asm. - if (ResultRegTypes.size() == 1) { + if (AsmInfo.ResultRegTypes.size() == 1) { RegResults.push_back(&Result); } else { - for (unsigned i = 0, e = ResultRegTypes.size(); i != e; ++i) { - llvm::Value *Tmp = CGF.Builder.CreateExtractValue(&Result, i, "asmresult"); + for (unsigned i = 0, e = AsmInfo.ResultRegTypes.size(); i != e; ++i) { + llvm::Value *Tmp = Builder.CreateExtractValue(&Result, i, "asmresult"); RegResults.push_back(Tmp); } } } -static void -EmitAsmStores(CodeGenFunction &CGF, const AsmStmt &S, - const llvm::ArrayRef<llvm::Value *> RegResults, - const llvm::ArrayRef<llvm::Type *> ResultRegTypes, - const llvm::ArrayRef<llvm::Type *> ResultTruncRegTypes, - const llvm::ArrayRef<LValue> ResultRegDests, - const llvm::ArrayRef<QualType> ResultRegQualTys, - const llvm::BitVector &ResultTypeRequiresCast, - const std::vector<std::optional<std::pair<unsigned, unsigned>>> - &ResultBounds) { - CGBuilderTy &Builder = CGF.Builder; - CodeGenModule &CGM = CGF.CGM; - llvm::LLVMContext &CTX = CGF.getLLVMContext(); - - assert(RegResults.size() == ResultRegTypes.size()); - assert(RegResults.size() == ResultTruncRegTypes.size()); - assert(RegResults.size() == ResultRegDests.size()); - // ResultRegDests can be also populated by addReturnRegisterOutputs() above, +void CodeGenFunction::EmitAsmStores( + const AsmStmt &S, const llvm::ArrayRef<llvm::Value *> RegResults, + const AsmConstraintsInfo &AsmInfo) { + llvm::LLVMContext &CTX = getLLVMContext(); + + assert(RegResults.size() == AsmInfo.ResultRegTypes.size()); + assert(RegResults.size() == AsmInfo.ResultTruncRegTypes.size()); + assert(RegResults.size() == AsmInfo.ResultRegDests.size()); + + // ResultRegDests can also be populated by addReturnRegisterOutputs() above, // in which case its size may grow. - assert(ResultTypeRequiresCast.size() <= ResultRegDests.size()); - assert(ResultBounds.size() <= ResultRegDests.size()); + assert(AsmInfo.ResultTypeRequiresCast.size() <= + AsmInfo.ResultRegDests.size()); + assert(AsmInfo.ResultBounds.size() <= AsmInfo.ResultRegDests.size()); for (unsigned i = 0, e = RegResults.size(); i != e; ++i) { llvm::Value *Tmp = RegResults[i]; - llvm::Type *TruncTy = ResultTruncRegTypes[i]; + llvm::Type *TruncTy = AsmInfo.ResultTruncRegTypes[i]; + + if (i < AsmInfo.ResultBounds.size() && + AsmInfo.ResultBounds[i].has_value()) { + const auto [LowerBound, UpperBound] = AsmInfo.ResultBounds[i].value(); - if ((i < ResultBounds.size()) && ResultBounds[i].has_value()) { - const auto [LowerBound, UpperBound] = ResultBounds[i].value(); // FIXME: Support for nonzero lower bounds not yet implemented. assert(LowerBound == 0 && "Output operand lower bound is not zero."); + llvm::Constant *UpperBoundConst = llvm::ConstantInt::get(Tmp->getType(), UpperBound); llvm::Value *IsBooleanValue = Builder.CreateCmp(llvm::CmpInst::ICMP_ULT, Tmp, UpperBoundConst); llvm::Function *FnAssume = CGM.getIntrinsic(llvm::Intrinsic::assume); + Builder.CreateCall(FnAssume, IsBooleanValue); } // If the result type of the LLVM IR asm doesn't match the result type of // the expression, do the conversion. - if (ResultRegTypes[i] != TruncTy) { - + if (AsmInfo.ResultRegTypes[i] != TruncTy) { // Truncate the integer result to the right size, note that TruncTy can be // a pointer. if (TruncTy->isFloatingPointTy()) @@ -2736,135 +2729,100 @@ EmitAsmStores(CodeGenFunction &CGF, const AsmStmt &S, } } - ApplyAtomGroup Grp(CGF.getDebugInfo()); - LValue Dest = ResultRegDests[i]; + ApplyAtomGroup Grp(getDebugInfo()); + LValue Dest = AsmInfo.ResultRegDests[i]; + // ResultTypeRequiresCast elements correspond to the first // ResultTypeRequiresCast.size() elements of RegResults. - if ((i < ResultTypeRequiresCast.size()) && ResultTypeRequiresCast[i]) { - unsigned Size = CGF.getContext().getTypeSize(ResultRegQualTys[i]); - Address A = Dest.getAddress().withElementType(ResultRegTypes[i]); - if (CGF.getTargetHooks().isScalarizableAsmOperand(CGF, TruncTy)) { + if (i < AsmInfo.ResultTypeRequiresCast.size() && + AsmInfo.ResultTypeRequiresCast[i]) { + unsigned Size = getContext().getTypeSize(AsmInfo.ResultRegQualTys[i]); + Address A = Dest.getAddress().withElementType(AsmInfo.ResultRegTypes[i]); + + if (getTargetHooks().isScalarizableAsmOperand(*this, TruncTy)) { llvm::StoreInst *S = Builder.CreateStore(Tmp, A); - CGF.addInstToCurrentSourceAtom(S, S->getValueOperand()); + addInstToCurrentSourceAtom(S, S->getValueOperand()); continue; } - QualType Ty = - CGF.getContext().getIntTypeForBitwidth(Size, /*Signed=*/false); + QualType Ty = getContext().getIntTypeForBitwidth(Size, /*Signed=*/false); if (Ty.isNull()) { const Expr *OutExpr = S.getOutputExpr(i); CGM.getDiags().Report(OutExpr->getExprLoc(), diag::err_store_value_to_reg); return; } - Dest = CGF.MakeAddrLValue(A, Ty); - } - CGF.EmitStoreThroughLValue(RValue::get(Tmp), Dest); - } -} - -static void EmitHipStdParUnsupportedAsm(CodeGenFunction *CGF, - const AsmStmt &S) { - constexpr auto Name = "__ASM__hipstdpar_unsupported"; - std::string Asm; - if (auto GCCAsm = dyn_cast<GCCAsmStmt>(&S)) - Asm = GCCAsm->getAsmString(); - - auto &Ctx = CGF->CGM.getLLVMContext(); - - auto StrTy = llvm::ConstantDataArray::getString(Ctx, Asm); - auto FnTy = llvm::FunctionType::get(llvm::Type::getVoidTy(Ctx), - {StrTy->getType()}, false); - auto UBF = CGF->CGM.getModule().getOrInsertFunction(Name, FnTy); + Dest = MakeAddrLValue(A, Ty); + } - CGF->Builder.CreateCall(UBF, {StrTy}); + EmitStoreThroughLValue(RValue::get(Tmp), Dest); + } } -void CodeGenFunction::EmitAsmStmt(const AsmStmt &S) { - // Pop all cleanup blocks at the end of the asm statement. - CodeGenFunction::RunCleanupsScope Cleanups(*this); - - // Assemble the final asm string. - std::string AsmString = S.generateAsmString(getContext()); - - // Get all the output and input constraints together. - SmallVector<TargetInfo::ConstraintInfo, 4> OutputConstraintInfos; - SmallVector<TargetInfo::ConstraintInfo, 4> InputConstraintInfos; - - bool IsHipStdPar = getLangOpts().HIPStdPar && getLangOpts().CUDAIsDevice; +/// Gather and validate the output and input constraints for the given inline +/// assembly statement. This ensures that the constraints are valid for the +/// target and prepares them for further processing. +bool CodeGenFunction::GetOutputAndInputConstraints( + const AsmStmt &S, + SmallVectorImpl<TargetInfo::ConstraintInfo> &OutputConstraintInfos, + SmallVectorImpl<TargetInfo::ConstraintInfo> &InputConstraintInfos) { bool IsValidTargetAsm = true; - for (unsigned i = 0, e = S.getNumOutputs(); i != e && IsValidTargetAsm; i++) { + bool IsHipStdPar = getLangOpts().HIPStdPar && getLangOpts().CUDAIsDevice; + for (unsigned I = 0, E = S.getNumOutputs(); I != E && IsValidTargetAsm; I++) { StringRef Name; if (const GCCAsmStmt *GAS = dyn_cast<GCCAsmStmt>(&S)) - Name = GAS->getOutputName(i); - TargetInfo::ConstraintInfo Info(S.getOutputConstraint(i), Name); - bool IsValid = getTarget().validateOutputConstraint(Info); (void)IsValid; + Name = GAS->getOutputName(I); + + TargetInfo::ConstraintInfo Info(S.getOutputConstraint(I), Name); + + bool IsValid = getTarget().validateOutputConstraint(Info); if (IsHipStdPar && !IsValid) IsValidTargetAsm = false; else assert(IsValid && "Failed to parse output constraint"); + OutputConstraintInfos.push_back(Info); } - for (unsigned i = 0, e = S.getNumInputs(); i != e && IsValidTargetAsm; i++) { + for (unsigned I = 0, E = S.getNumInputs(); I != E && IsValidTargetAsm; I++) { StringRef Name; if (const GCCAsmStmt *GAS = dyn_cast<GCCAsmStmt>(&S)) - Name = GAS->getInputName(i); - TargetInfo::ConstraintInfo Info(S.getInputConstraint(i), Name); + Name = GAS->getInputName(I); + + TargetInfo::ConstraintInfo Info(S.getInputConstraint(I), Name); + bool IsValid = - getTarget().validateInputConstraint(OutputConstraintInfos, Info); + getTarget().validateInputConstraint(OutputConstraintInfos, Info); if (IsHipStdPar && !IsValid) IsValidTargetAsm = false; else assert(IsValid && "Failed to parse input constraint"); + InputConstraintInfos.push_back(Info); } - if (!IsValidTargetAsm) - return EmitHipStdParUnsupportedAsm(this, S); - - std::string Constraints; - - std::vector<LValue> ResultRegDests; - std::vector<QualType> ResultRegQualTys; - std::vector<llvm::Type *> ResultRegTypes; - std::vector<llvm::Type *> ResultTruncRegTypes; - std::vector<llvm::Type *> ArgTypes; - std::vector<llvm::Type *> ArgElemTypes; - std::vector<llvm::Value*> Args; - llvm::BitVector ResultTypeRequiresCast; - std::vector<std::optional<std::pair<unsigned, unsigned>>> ResultBounds; - - // Keep track of inout constraints. - std::string InOutConstraints; - std::vector<llvm::Value*> InOutArgs; - std::vector<llvm::Type*> InOutArgTypes; - std::vector<llvm::Type*> InOutArgElemTypes; - - // Keep track of out constraints for tied input operand. - std::vector<std::string> OutputConstraints; + return IsValidTargetAsm; +} +/// Process the output constraints of an inline assembly statement. This method +/// handles the complexity of determining whether an output should be a +/// register or memory operand, manages tied operands, and prepares the +/// necessary arguments for the LLVM inline asm call. +void CodeGenFunction::HandleOutputConstraints(const AsmStmt &S, + AsmConstraintsInfo &AsmInfo) { // Keep track of defined physregs. llvm::SmallSet<std::string, 8> PhysRegOutputs; - // An inline asm can be marked readonly if it meets the following conditions: - // - it doesn't have any sideeffects - // - it doesn't clobber memory - // - it doesn't return a value by-reference - // It can be marked readnone if it doesn't have any input memory constraints - // in addition to meeting the conditions listed above. - bool ReadOnly = true, ReadNone = true; - - for (unsigned i = 0, e = S.getNumOutputs(); i != e; i++) { - TargetInfo::ConstraintInfo &Info = OutputConstraintInfos[i]; + for (unsigned I = 0, E = S.getNumOutputs(); I != E; I++) { + TargetInfo::ConstraintInfo &Info = AsmInfo.OutputConstraintInfos[I]; // Simplify the output constraint. - std::string OutputConstraint(S.getOutputConstraint(i)); + std::string OutputConstraint(S.getOutputConstraint(I)); OutputConstraint = getTarget().simplifyConstraint( - StringRef(OutputConstraint).substr(1), &OutputConstraintInfos); + StringRef(OutputConstraint).substr(1), &AsmInfo.OutputConstraintInfos); - const Expr *OutExpr = S.getOutputExpr(i); + const Expr *OutExpr = S.getOutputExpr(I); OutExpr = OutExpr->IgnoreParenNoopCasts(getContext()); std::string GCCReg; @@ -2874,52 +2832,56 @@ void CodeGenFunction::EmitAsmStmt(const AsmStmt &S) { CGM.ErrorUnsupported(UnspStmt, Msg); }, &GCCReg); + // Give an error on multiple outputs to same physreg. if (!GCCReg.empty() && !PhysRegOutputs.insert(GCCReg).second) CGM.Error(S.getAsmLoc(), "multiple outputs to hard register: " + GCCReg); - OutputConstraints.push_back(OutputConstraint); + AsmInfo.OutputConstraints.push_back(OutputConstraint); LValue Dest = EmitLValue(OutExpr); - if (!Constraints.empty()) - Constraints += ','; + if (!AsmInfo.Constraints.empty()) + AsmInfo.Constraints += ','; // If this is a register output, then make the inline asm return it // by-value. If this is a memory result, return the value by-reference. QualType QTy = OutExpr->getType(); - const bool IsScalarOrAggregate = hasScalarEvaluationKind(QTy) || - hasAggregateEvaluationKind(QTy); - if (!Info.allowsMemory() && IsScalarOrAggregate) { + const bool IsScalarOrAggregate = + hasScalarEvaluationKind(QTy) || hasAggregateEvaluationKind(QTy); - Constraints += "=" + OutputConstraint; - ResultRegQualTys.push_back(QTy); - ResultRegDests.push_back(Dest); + if (!Info.allowsMemory() && IsScalarOrAggregate) { + AsmInfo.Constraints += "=" + OutputConstraint; + AsmInfo.ResultRegQualTys.push_back(QTy); + AsmInfo.ResultRegDests.push_back(Dest); - ResultBounds.emplace_back(Info.getOutputOperandBounds()); + AsmInfo.ResultBounds.emplace_back(Info.getOutputOperandBounds()); llvm::Type *Ty = ConvertTypeForMem(QTy); - const bool RequiresCast = Info.allowsRegister() && + const bool RequiresCast = + Info.allowsRegister() && (getTargetHooks().isScalarizableAsmOperand(*this, Ty) || Ty->isAggregateType()); - ResultTruncRegTypes.push_back(Ty); - ResultTypeRequiresCast.push_back(RequiresCast); + AsmInfo.ResultTruncRegTypes.push_back(Ty); + AsmInfo.ResultTypeRequiresCast.push_back(RequiresCast); if (RequiresCast) { - unsigned Size = getContext().getTypeSize(QTy); - if (Size) + if (unsigned Size = getContext().getTypeSize(QTy)) Ty = llvm::IntegerType::get(getLLVMContext(), Size); else CGM.Error(OutExpr->getExprLoc(), "output size should not be zero"); } - ResultRegTypes.push_back(Ty); + + AsmInfo.ResultRegTypes.push_back(Ty); + // If this output is tied to an input, and if the input is larger, then // we need to set the actual result type of the inline asm node to be the // same as the input type. if (Info.hasMatchingInput()) { unsigned InputNo; for (InputNo = 0; InputNo != S.getNumInputs(); ++InputNo) { - TargetInfo::ConstraintInfo &Input = InputConstraintInfos[InputNo]; - if (Input.hasTiedOperand() && Input.getTiedOperand() == i) + TargetInfo::ConstraintInfo &Input = + AsmInfo.InputConstraintInfos[InputNo]; + if (Input.hasTiedOperand() && Input.getTiedOperand() == I) break; } assert(InputNo != S.getNumInputs() && "Didn't find matching input!"); @@ -2928,28 +2890,27 @@ void CodeGenFunction::EmitAsmStmt(const AsmStmt &S) { QualType OutputType = OutExpr->getType(); uint64_t InputSize = getContext().getTypeSize(InputTy); - if (getContext().getTypeSize(OutputType) < InputSize) { + if (getContext().getTypeSize(OutputType) < InputSize) // Form the asm to return the value as a larger integer or fp type. - ResultRegTypes.back() = ConvertType(InputTy); - } + AsmInfo.ResultRegTypes.back() = ConvertType(InputTy); } - if (llvm::Type* AdjTy = - getTargetHooks().adjustInlineAsmType(*this, OutputConstraint, - ResultRegTypes.back())) - ResultRegTypes.back() = AdjTy; - else { + + if (llvm::Type *AdjTy = getTargetHooks().adjustInlineAsmType( + *this, OutputConstraint, AsmInfo.ResultRegTypes.back())) + AsmInfo.ResultRegTypes.back() = AdjTy; + else CGM.getDiags().Report(S.getAsmLoc(), diag::err_asm_invalid_type_in_input) << OutExpr->getType() << OutputConstraint; - } // Update largest vector width for any vector types. - if (auto *VT = dyn_cast<llvm::VectorType>(ResultRegTypes.back())) + if (auto *VT = dyn_cast<llvm::VectorType>(AsmInfo.ResultRegTypes.back())) LargestVectorWidth = std::max((uint64_t)LargestVectorWidth, VT->getPrimitiveSizeInBits().getKnownMinValue()); } else { Address DestAddr = Dest.getAddress(); + // Matrix types in memory are represented by arrays, but accessed through // vector pointers, with the alignment specified on the access operation. // For inline assembly, update pointer arguments to use vector pointers. @@ -2958,87 +2919,105 @@ void CodeGenFunction::EmitAsmStmt(const AsmStmt &S) { if (isa<MatrixType>(OutExpr->getType().getCanonicalType())) DestAddr = DestAddr.withElementType(ConvertType(OutExpr->getType())); - ArgTypes.push_back(DestAddr.getType()); - ArgElemTypes.push_back(DestA... [truncated] `````````` </details> https://github.com/llvm/llvm-project/pull/196885 _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
