Skip to content

Commit 8baa770

Browse files
piyushjaiswal98piyushjaiswal98
authored
Merge branch 'main' into breakpoint_change_notif
2 parents e04bbb2 + 9b5bc98 commit 8baa770

File tree

75 files changed

+2143
-466
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

75 files changed

+2143
-466
lines changed

‎clang/include/clang/Basic/BuiltinsAMDGPU.def‎

Lines changed: 9 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -830,6 +830,15 @@ TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b4_u4, "V2UiUiUiV2Ui", "nc", "tensor-c
830830
TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b6_u4, "V3UiUiULiV2Ui", "nc", "tensor-cvt-lut-insts")
831831
TARGET_BUILTIN(__builtin_amdgcn_perm_pk16_b8_u4, "V4UiULiULiV2Ui", "nc", "tensor-cvt-lut-insts")
832832

833+
TARGET_BUILTIN(__builtin_amdgcn_add_max_i32, "iiiiIb", "nc", "add-min-max-insts")
834+
TARGET_BUILTIN(__builtin_amdgcn_add_max_u32, "UiUiUiUiIb", "nc", "add-min-max-insts")
835+
TARGET_BUILTIN(__builtin_amdgcn_add_min_i32, "iiiiIb", "nc", "add-min-max-insts")
836+
TARGET_BUILTIN(__builtin_amdgcn_add_min_u32, "UiUiUiUiIb", "nc", "add-min-max-insts")
837+
TARGET_BUILTIN(__builtin_amdgcn_pk_add_max_i16, "V2sV2sV2sV2sIb", "nc", "pk-add-min-max-insts")
838+
TARGET_BUILTIN(__builtin_amdgcn_pk_add_max_u16, "V2UsV2UsV2UsV2UsIb", "nc", "pk-add-min-max-insts")
839+
TARGET_BUILTIN(__builtin_amdgcn_pk_add_min_i16, "V2sV2sV2sV2sIb", "nc", "pk-add-min-max-insts")
840+
TARGET_BUILTIN(__builtin_amdgcn_pk_add_min_u16, "V2UsV2UsV2UsV2UsIb", "nc", "pk-add-min-max-insts")
841+
833842
// GFX1250 WMMA builtins
834843
TARGET_BUILTIN(__builtin_amdgcn_wmma_f32_16x16x4_f32, "V8fIbV2fIbV2fIsV8fIbIb", "nc", "gfx1250-insts,wavefrontsize32")
835844
TARGET_BUILTIN(__builtin_amdgcn_wmma_f32_16x16x32_bf16, "V8fIbV16yIbV16yIsV8fIbIb", "nc", "gfx1250-insts,wavefrontsize32")

‎clang/lib/Analysis/FlowSensitive/Models/UncheckedStatusOrAccessModel.cpp‎

Lines changed: 37 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -150,6 +150,24 @@ static auto isComparisonOperatorCall(llvm::StringRef operator_name) {
150150
hasArgument(1, anyOf(hasType(statusType()), hasType(statusOrType()))));
151151
}
152152

153+
static auto isOkStatusCall() {
154+
using namespace ::clang::ast_matchers; // NOLINT: Too many names
155+
return callExpr(callee(functionDecl(hasName("::absl::OkStatus"))));
156+
}
157+
158+
static auto isNotOkStatusCall() {
159+
using namespace ::clang::ast_matchers; // NOLINT: Too many names
160+
return callExpr(callee(functionDecl(hasAnyName(
161+
"::absl::AbortedError", "::absl::AlreadyExistsError",
162+
"::absl::CancelledError", "::absl::DataLossError",
163+
"::absl::DeadlineExceededError", "::absl::FailedPreconditionError",
164+
"::absl::InternalError", "::absl::InvalidArgumentError",
165+
"::absl::NotFoundError", "::absl::OutOfRangeError",
166+
"::absl::PermissionDeniedError", "::absl::ResourceExhaustedError",
167+
"::absl::UnauthenticatedError", "::absl::UnavailableError",
168+
"::absl::UnimplementedError", "::absl::UnknownError"))));
169+
}
170+
153171
static auto
154172
buildDiagnoseMatchSwitch(const UncheckedStatusOrAccessModelOptions &Options) {
155173
return CFGMatchSwitchBuilder<const Environment,
@@ -420,6 +438,23 @@ static void transferComparisonOperator(const CXXOperatorCallExpr *Expr,
420438
State.Env.setValue(*Expr, *LhsAndRhsVal);
421439
}
422440

441+
static void transferOkStatusCall(const CallExpr *Expr,
442+
const MatchFinder::MatchResult &,
443+
LatticeTransferState &State) {
444+
auto &OkVal =
445+
initializeStatus(State.Env.getResultObjectLocation(*Expr), State.Env);
446+
State.Env.assume(OkVal.formula());
447+
}
448+
449+
static void transferNotOkStatusCall(const CallExpr *Expr,
450+
const MatchFinder::MatchResult &,
451+
LatticeTransferState &State) {
452+
auto &OkVal =
453+
initializeStatus(State.Env.getResultObjectLocation(*Expr), State.Env);
454+
auto &A = State.Env.arena();
455+
State.Env.assume(A.makeNot(OkVal.formula()));
456+
}
457+
423458
CFGMatchSwitch<LatticeTransferState>
424459
buildTransferMatchSwitch(ASTContext &Ctx,
425460
CFGMatchSwitchBuilder<LatticeTransferState> Builder) {
@@ -447,6 +482,8 @@ buildTransferMatchSwitch(ASTContext &Ctx,
447482
transferComparisonOperator(Expr, State,
448483
/*IsNegative=*/true);
449484
})
485+
.CaseOfCFGStmt<CallExpr>(isOkStatusCall(), transferOkStatusCall)
486+
.CaseOfCFGStmt<CallExpr>(isNotOkStatusCall(), transferNotOkStatusCall)
450487
.Build();
451488
}
452489

‎clang/lib/CIR/CodeGen/CIRGenBuilder.cpp‎

Lines changed: 5 additions & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -22,8 +22,8 @@ mlir::Value CIRGenBuilderTy::maybeBuildArrayDecay(mlir::Location loc,
2222

2323
if (arrayTy) {
2424
const cir::PointerType flatPtrTy = getPointerTo(arrayTy.getElementType());
25-
return create<cir::CastOp>(loc, flatPtrTy, cir::CastKind::array_to_ptrdecay,
26-
arrayPtr);
25+
return cir::CastOp::create(*this, loc, flatPtrTy,
26+
cir::CastKind::array_to_ptrdecay, arrayPtr);
2727
}
2828

2929
assert(arrayPtrTy.getPointee() == eltTy &&
@@ -40,7 +40,7 @@ mlir::Value CIRGenBuilderTy::getArrayElement(mlir::Location arrayLocBegin,
4040
if (shouldDecay)
4141
basePtr = maybeBuildArrayDecay(arrayLocBegin, arrayPtr, eltTy);
4242
const mlir::Type flatPtrTy = basePtr.getType();
43-
return create<cir::PtrStrideOp>(arrayLocEnd, flatPtrTy, basePtr, idx);
43+
return cir::PtrStrideOp::create(*this, arrayLocEnd, flatPtrTy, basePtr, idx);
4444
}
4545

4646
cir::ConstantOp CIRGenBuilderTy::getConstInt(mlir::Location loc,
@@ -60,14 +60,14 @@ cir::ConstantOp CIRGenBuilderTy::getConstInt(mlir::Location loc,
6060
cir::ConstantOp CIRGenBuilderTy::getConstInt(mlir::Location loc, mlir::Type t,
6161
uint64_t c) {
6262
assert(mlir::isa<cir::IntType>(t) && "expected cir::IntType");
63-
return create<cir::ConstantOp>(loc, cir::IntAttr::get(t, c));
63+
return cir::ConstantOp::create(*this, loc, cir::IntAttr::get(t, c));
6464
}
6565

6666
cir::ConstantOp
6767
clang::CIRGen::CIRGenBuilderTy::getConstFP(mlir::Location loc, mlir::Type t,
6868
llvm::APFloat fpVal) {
6969
assert(mlir::isa<cir::FPTypeInterface>(t) && "expected floating point type");
70-
return create<cir::ConstantOp>(loc, cir::FPAttr::get(t, fpVal));
70+
return cir::ConstantOp::create(*this, loc, cir::FPAttr::get(t, fpVal));
7171
}
7272

7373
void CIRGenBuilderTy::computeGlobalViewIndicesFromFlatOffset(

‎clang/lib/CIR/CodeGen/CIRGenCXXABI.cpp‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -70,8 +70,8 @@ cir::GlobalLinkageKind CIRGenCXXABI::getCXXDestructorLinkage(
7070
mlir::Value CIRGenCXXABI::loadIncomingCXXThis(CIRGenFunction &cgf) {
7171
ImplicitParamDecl *vd = getThisDecl(cgf);
7272
Address addr = cgf.getAddrOfLocalVar(vd);
73-
return cgf.getBuilder().create<cir::LoadOp>(
74-
cgf.getLoc(vd->getLocation()), addr.getElementType(), addr.getPointer());
73+
return cir::LoadOp::create(cgf.getBuilder(), cgf.getLoc(vd->getLocation()),
74+
addr.getElementType(), addr.getPointer());
7575
}
7676

7777
void CIRGenCXXABI::setCXXABIThisValue(CIRGenFunction &cgf,

‎clang/lib/CIR/CodeGen/CIRGenCall.cpp‎

Lines changed: 16 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -690,6 +690,22 @@ void CallArg::copyInto(CIRGenFunction &cgf, Address addr,
690690
isUsed = true;
691691
}
692692

693+
mlir::Value CIRGenFunction::emitRuntimeCall(mlir::Location loc,
694+
cir::FuncOp callee,
695+
ArrayRef<mlir::Value> args) {
696+
// TODO(cir): set the calling convention to this runtime call.
697+
assert(!cir::MissingFeatures::opFuncCallingConv());
698+
699+
cir::CallOp call = builder.createCallOp(loc, callee, args);
700+
assert(call->getNumResults() <= 1 &&
701+
"runtime functions have at most 1 result");
702+
703+
if (call->getNumResults() == 0)
704+
return nullptr;
705+
706+
return call->getResult(0);
707+
}
708+
693709
void CIRGenFunction::emitCallArg(CallArgList &args, const clang::Expr *e,
694710
clang::QualType argType) {
695711
assert(argType->isReferenceType() == e->isGLValue() &&

‎clang/lib/CIR/CodeGen/CIRGenExpr.cpp‎

Lines changed: 19 additions & 18 deletions
Original file line numberDiff line numberDiff line change
@@ -671,8 +671,8 @@ static LValue emitFunctionDeclLValue(CIRGenFunction &cgf, const Expr *e,
671671

672672
mlir::Type fnTy = funcOp.getFunctionType();
673673
mlir::Type ptrTy = cir::PointerType::get(fnTy);
674-
mlir::Value addr = cgf.getBuilder().create<cir::GetGlobalOp>(
675-
loc, ptrTy, funcOp.getSymName());
674+
mlir::Value addr = cir::GetGlobalOp::create(cgf.getBuilder(), loc, ptrTy,
675+
funcOp.getSymName());
676676

677677
if (funcOp.getFunctionType() != cgf.convertType(fd->getType())) {
678678
fnTy = cgf.convertType(fd->getType());
@@ -1820,10 +1820,12 @@ CIRGenCallee CIRGenFunction::emitCallee(const clang::Expr *e) {
18201820
// Resolve direct calls.
18211821
const auto *funcDecl = cast<FunctionDecl>(declRef->getDecl());
18221822
return emitDirectCallee(funcDecl);
1823-
} else if (isa<MemberExpr>(e)) {
1824-
cgm.errorNYI(e->getSourceRange(),
1825-
"emitCallee: call to member function is NYI");
1826-
return {};
1823+
} else if (auto me = dyn_cast<MemberExpr>(e)) {
1824+
if (const auto *fd = dyn_cast<FunctionDecl>(me->getMemberDecl())) {
1825+
emitIgnoredExpr(me->getBase());
1826+
return emitDirectCallee(fd);
1827+
}
1828+
// Else fall through to the indirect reference handling below.
18271829
} else if (auto *pde = dyn_cast<CXXPseudoDestructorExpr>(e)) {
18281830
return CIRGenCallee::forPseudoDestructor(pde);
18291831
}
@@ -2020,18 +2022,17 @@ mlir::Value CIRGenFunction::emitOpOnBoolExpr(mlir::Location loc,
20202022
mlir::Value condV = emitOpOnBoolExpr(loc, condOp->getCond());
20212023

20222024
mlir::Value ternaryOpRes =
2023-
builder
2024-
.create<cir::TernaryOp>(
2025-
loc, condV, /*thenBuilder=*/
2026-
[this, trueExpr](mlir::OpBuilder &b, mlir::Location loc) {
2027-
mlir::Value lhs = emitScalarExpr(trueExpr);
2028-
cir::YieldOp::create(b, loc, lhs);
2029-
},
2030-
/*elseBuilder=*/
2031-
[this, falseExpr](mlir::OpBuilder &b, mlir::Location loc) {
2032-
mlir::Value rhs = emitScalarExpr(falseExpr);
2033-
cir::YieldOp::create(b, loc, rhs);
2034-
})
2025+
cir::TernaryOp::create(
2026+
builder, loc, condV, /*thenBuilder=*/
2027+
[this, trueExpr](mlir::OpBuilder &b, mlir::Location loc) {
2028+
mlir::Value lhs = emitScalarExpr(trueExpr);
2029+
cir::YieldOp::create(b, loc, lhs);
2030+
},
2031+
/*elseBuilder=*/
2032+
[this, falseExpr](mlir::OpBuilder &b, mlir::Location loc) {
2033+
mlir::Value rhs = emitScalarExpr(falseExpr);
2034+
cir::YieldOp::create(b, loc, rhs);
2035+
})
20352036
.getResult();
20362037

20372038
return emitScalarConversion(ternaryOpRes, condOp->getType(),

‎clang/lib/CIR/CodeGen/CIRGenExprComplex.cpp‎

Lines changed: 16 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -969,23 +969,22 @@ mlir::Value ComplexExprEmitter::VisitAbstractConditionalOperator(
969969
Expr *cond = e->getCond()->IgnoreParens();
970970
mlir::Value condValue = cgf.evaluateExprAsBool(cond);
971971

972-
return builder
973-
.create<cir::TernaryOp>(
974-
loc, condValue,
975-
/*thenBuilder=*/
976-
[&](mlir::OpBuilder &b, mlir::Location loc) {
977-
eval.beginEvaluation();
978-
mlir::Value trueValue = Visit(e->getTrueExpr());
979-
cir::YieldOp::create(b, loc, trueValue);
980-
eval.endEvaluation();
981-
},
982-
/*elseBuilder=*/
983-
[&](mlir::OpBuilder &b, mlir::Location loc) {
984-
eval.beginEvaluation();
985-
mlir::Value falseValue = Visit(e->getFalseExpr());
986-
cir::YieldOp::create(b, loc, falseValue);
987-
eval.endEvaluation();
988-
})
972+
return cir::TernaryOp::create(
973+
builder, loc, condValue,
974+
/*thenBuilder=*/
975+
[&](mlir::OpBuilder &b, mlir::Location loc) {
976+
eval.beginEvaluation();
977+
mlir::Value trueValue = Visit(e->getTrueExpr());
978+
cir::YieldOp::create(b, loc, trueValue);
979+
eval.endEvaluation();
980+
},
981+
/*elseBuilder=*/
982+
[&](mlir::OpBuilder &b, mlir::Location loc) {
983+
eval.beginEvaluation();
984+
mlir::Value falseValue = Visit(e->getFalseExpr());
985+
cir::YieldOp::create(b, loc, falseValue);
986+
eval.endEvaluation();
987+
})
989988
.getResult();
990989
}
991990

‎clang/lib/CIR/CodeGen/CIRGenExprConstant.cpp‎

Lines changed: 17 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -179,8 +179,23 @@ bool ConstantAggregateBuilder::add(mlir::TypedAttr typedAttr, CharUnits offset,
179179
}
180180

181181
// Uncommon case: constant overlaps what we've already created.
182-
cgm.errorNYI("overlapping constants");
183-
return false;
182+
std::optional<size_t> firstElemToReplace = splitAt(offset);
183+
if (!firstElemToReplace)
184+
return false;
185+
186+
CharUnits cSize = getSize(typedAttr);
187+
std::optional<size_t> lastElemToReplace = splitAt(offset + cSize);
188+
if (!lastElemToReplace)
189+
return false;
190+
191+
assert((firstElemToReplace == lastElemToReplace || allowOverwrite) &&
192+
"unexpectedly overwriting field");
193+
194+
Element newElt(typedAttr, offset);
195+
replace(elements, *firstElemToReplace, *lastElemToReplace, {newElt});
196+
size = std::max(size, offset + cSize);
197+
naturalLayout = false;
198+
return true;
184199
}
185200

186201
bool ConstantAggregateBuilder::addBits(llvm::APInt bits, uint64_t offsetInBits,

‎clang/lib/CIR/CodeGen/CIRGenExprScalar.cpp‎

Lines changed: 16 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -1568,8 +1568,9 @@ static mlir::Value emitPointerArithmetic(CIRGenFunction &cgf,
15681568
}
15691569

15701570
assert(!cir::MissingFeatures::sanitizers());
1571-
return cgf.getBuilder().create<cir::PtrStrideOp>(
1572-
cgf.getLoc(op.e->getExprLoc()), pointer.getType(), pointer, index);
1571+
return cir::PtrStrideOp::create(cgf.getBuilder(),
1572+
cgf.getLoc(op.e->getExprLoc()),
1573+
pointer.getType(), pointer, index);
15731574
}
15741575

15751576
mlir::Value ScalarExprEmitter::emitMul(const BinOpInfo &ops) {
@@ -2075,8 +2076,9 @@ mlir::Value ScalarExprEmitter::VisitInitListExpr(InitListExpr *e) {
20752076
vectorType.getSize() - numInitElements, zeroValue);
20762077
}
20772078

2078-
return cgf.getBuilder().create<cir::VecCreateOp>(
2079-
cgf.getLoc(e->getSourceRange()), vectorType, elements);
2079+
return cir::VecCreateOp::create(cgf.getBuilder(),
2080+
cgf.getLoc(e->getSourceRange()), vectorType,
2081+
elements);
20802082
}
20812083

20822084
// C++11 value-initialization for the scalar.
@@ -2364,17 +2366,16 @@ mlir::Value ScalarExprEmitter::VisitAbstractConditionalOperator(
23642366
}
23652367
};
23662368

2367-
mlir::Value result = builder
2368-
.create<cir::TernaryOp>(
2369-
loc, condV,
2370-
/*trueBuilder=*/
2371-
[&](mlir::OpBuilder &b, mlir::Location loc) {
2372-
emitBranch(b, loc, lhsExpr);
2373-
},
2374-
/*falseBuilder=*/
2375-
[&](mlir::OpBuilder &b, mlir::Location loc) {
2376-
emitBranch(b, loc, rhsExpr);
2377-
})
2369+
mlir::Value result = cir::TernaryOp::create(
2370+
builder, loc, condV,
2371+
/*trueBuilder=*/
2372+
[&](mlir::OpBuilder &b, mlir::Location loc) {
2373+
emitBranch(b, loc, lhsExpr);
2374+
},
2375+
/*falseBuilder=*/
2376+
[&](mlir::OpBuilder &b, mlir::Location loc) {
2377+
emitBranch(b, loc, rhsExpr);
2378+
})
23782379
.getResult();
23792380

23802381
if (!insertPoints.empty()) {

‎clang/lib/CIR/CodeGen/CIRGenFunction.h‎

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1461,6 +1461,9 @@ class CIRGenFunction : public CIRGenTypeCache {
14611461

14621462
void emitReturnOfRValue(mlir::Location loc, RValue rv, QualType ty);
14631463

1464+
mlir::Value emitRuntimeCall(mlir::Location loc, cir::FuncOp callee,
1465+
llvm::ArrayRef<mlir::Value> args = {});
1466+
14641467
/// Emit the computation of the specified expression of scalar type.
14651468
mlir::Value emitScalarExpr(const clang::Expr *e);
14661469

0 commit comments

Comments
 (0)