Integrate LLVM at llvm/llvm-project@6706a4720ffc
Updates LLVM usage to match
[6706a4720ffc](https://github.com/llvm/llvm-project/commit/6706a4720ffc)
PiperOrigin-RevId: 408566046
diff --git a/SUBMODULE_VERSIONS.txt b/SUBMODULE_VERSIONS.txt
index fe0fb7d..8f32fd1 100644
--- a/SUBMODULE_VERSIONS.txt
+++ b/SUBMODULE_VERSIONS.txt
@@ -4,7 +4,7 @@
aa533abfd4232b01f9e57041d70114d5a77e6de0 third_party/googletest
88b845dee001723c4a0db1fe5477de735b6d3bb0 third_party/liburing
acd6f6f014c25e46363e718381e0b35205df2d83 third_party/libyaml
-ca47447952f1f8b0de11aac75b45f83f88579b80 third_party/llvm-project
+6706a4720ffc229eb0b606728b26e988a98c6262 third_party/llvm-project
824581d8e6898bc12c2890b390a6b781d29303e7 third_party/mlir-hlo
3f701faace7addc75d16dea8a6cd769fa5b3f260 third_party/musl
4c7697dbe973ed01ae6fbec37d186ebd05982e1f third_party/pybind11
diff --git a/iree/compiler/Codegen/Common/VectorizeConv.cpp b/iree/compiler/Codegen/Common/VectorizeConv.cpp
index c976f5c..11b6736 100644
--- a/iree/compiler/Codegen/Common/VectorizeConv.cpp
+++ b/iree/compiler/Codegen/Common/VectorizeConv.cpp
@@ -97,8 +97,8 @@
int64_t numOutputHeights = outputShape[1];
int64_t numOutputWidths = outputShape[2];
- int64_t heightStride = convOp.strides().getValue<int64_t>({0});
- int64_t widthStride = convOp.strides().getValue<int64_t>({1});
+ int64_t heightStride = convOp.strides().getValues<int64_t>()[0];
+ int64_t widthStride = convOp.strides().getValues<int64_t>()[1];
// This invocation handles a batch of
// (numOutputHeights * numOutputWidths * numOutputChannels).
@@ -274,8 +274,8 @@
int64_t numOutputHeights = outputShape[1];
int64_t numOutputWidths = outputShape[2];
- int64_t heightStride = convOp.strides().getValue<int64_t>({0});
- int64_t widthStride = convOp.strides().getValue<int64_t>({1});
+ int64_t heightStride = convOp.strides().getValues<int64_t>()[0];
+ int64_t widthStride = convOp.strides().getValues<int64_t>()[1];
// This invocation handles a batch of (numOutputHeights * numOutputWidths *
// numChannels).
diff --git a/iree/compiler/Codegen/SPIRV/SPIRVVectorToCooperativeOps.cpp b/iree/compiler/Codegen/SPIRV/SPIRVVectorToCooperativeOps.cpp
index 170f1bc..992bba2 100644
--- a/iree/compiler/Codegen/SPIRV/SPIRVVectorToCooperativeOps.cpp
+++ b/iree/compiler/Codegen/SPIRV/SPIRVVectorToCooperativeOps.cpp
@@ -154,7 +154,7 @@
// Only convert splat integer/float vectors.
auto values = op.value().dyn_cast<DenseIntOrFPElementsAttr>();
if (!values || !values.isSplat()) return failure();
- Attribute value = values.getSplatValue();
+ Attribute value = values.getSplatValue<Attribute>();
auto elementType = values.getType().getElementType();
Value splatValue = rewriter.create<spirv::ConstantOp>(
diff --git a/iree/compiler/Dialect/Flow/IR/FlowOpFolders.cpp b/iree/compiler/Dialect/Flow/IR/FlowOpFolders.cpp
index ef12029..463c1d7 100644
--- a/iree/compiler/Dialect/Flow/IR/FlowOpFolders.cpp
+++ b/iree/compiler/Dialect/Flow/IR/FlowOpFolders.cpp
@@ -731,10 +731,10 @@
// Load directly from the constant source tensor.
auto indices = operands.drop_front();
if (llvm::count(indices, nullptr) == 0) {
- return source.getValue(
- llvm::to_vector<4>(llvm::map_range(indices, [](Attribute value) {
+ return source.getValues<Attribute>()[llvm::to_vector<4>(
+ llvm::map_range(indices, [](Attribute value) {
return value.cast<IntegerAttr>().getValue().getZExtValue();
- })));
+ }))];
}
}
return {};
@@ -864,7 +864,7 @@
int64_t numElements = updateType.getNumElements();
while (numElements--) {
targetValues[getFlattenedIndex(targetType, targetIndex)] =
- update.getValue<Attribute>(updateIndex);
+ update.getValues<Attribute>()[updateIndex];
// Increment the index at last dim.
++updateIndex.back();
++targetIndex.back();
diff --git a/iree/compiler/Dialect/Flow/Transforms/ConvertConv2DToImg2ColPass.cpp b/iree/compiler/Dialect/Flow/Transforms/ConvertConv2DToImg2ColPass.cpp
index cf64a53..92b7f63 100644
--- a/iree/compiler/Dialect/Flow/Transforms/ConvertConv2DToImg2ColPass.cpp
+++ b/iree/compiler/Dialect/Flow/Transforms/ConvertConv2DToImg2ColPass.cpp
@@ -98,7 +98,7 @@
auto s = [&](unsigned i) {
return rewriter.getAffineConstantExpr(
- convOp.strides().getValue<int64_t>({i}));
+ convOp.strides().getValues<int64_t>()[i]);
};
SmallVector<AffineExpr, 4> inputExprs = {n, d(1) * s(0) + k(1),
@@ -263,7 +263,7 @@
auto s = [&](unsigned i) {
return rewriter.getAffineConstantExpr(
- convOp.strides().getValue<int64_t>({i}));
+ convOp.strides().getValues<int64_t>()[i]);
};
SmallVector<AffineExpr> inputExprs = {n, ci, d(1) * s(0) + k(1),
diff --git a/iree/compiler/Dialect/Flow/Transforms/ConvertLinalgTensorOps.cpp b/iree/compiler/Dialect/Flow/Transforms/ConvertLinalgTensorOps.cpp
index 992878b..3e09e34 100644
--- a/iree/compiler/Dialect/Flow/Transforms/ConvertLinalgTensorOps.cpp
+++ b/iree/compiler/Dialect/Flow/Transforms/ConvertLinalgTensorOps.cpp
@@ -102,7 +102,8 @@
}
auto tensorType = op.getType().cast<TensorType>();
auto elementValue = rewriter.createOrFold<mlir::ConstantOp>(
- op.getLoc(), tensorType.getElementType(), splatAttr.getSplatValue());
+ op.getLoc(), tensorType.getElementType(),
+ splatAttr.getSplatValue<Attribute>());
rewriter.replaceOpWithNewOp<IREE::Flow::TensorSplatOp>(
op, tensorType, elementValue, ValueRange{});
return success();
diff --git a/iree/compiler/Dialect/HAL/Transforms/MaterializeConstantPoolBuffers.cpp b/iree/compiler/Dialect/HAL/Transforms/MaterializeConstantPoolBuffers.cpp
index ffa9f2a..eb506a6 100644
--- a/iree/compiler/Dialect/HAL/Transforms/MaterializeConstantPoolBuffers.cpp
+++ b/iree/compiler/Dialect/HAL/Transforms/MaterializeConstantPoolBuffers.cpp
@@ -265,7 +265,7 @@
auto lengthValue = builder.createOrFold<mlir::arith::ConstantIndexOp>(
splatOp.getLoc(), runtimeRange.getLength());
uint32_t pattern = makePatternFromSplatValue(
- splatOp.value().cast<SplatElementsAttr>().getSplatValue());
+ splatOp.value().cast<SplatElementsAttr>().getSplatValue<Attribute>());
auto patternValue = builder.createOrFold<mlir::arith::ConstantIntOp>(
loc, static_cast<int64_t>(pattern), 32);
builder.create<IREE::HAL::CommandBufferFillBufferOp>(
diff --git a/iree/compiler/Dialect/Stream/IR/StreamOpFolders.cpp b/iree/compiler/Dialect/Stream/IR/StreamOpFolders.cpp
index 473df80..e663b0d 100644
--- a/iree/compiler/Dialect/Stream/IR/StreamOpFolders.cpp
+++ b/iree/compiler/Dialect/Stream/IR/StreamOpFolders.cpp
@@ -683,7 +683,7 @@
"only constant splat attrs can be converted to splat ops");
}
- auto splatElementAttr = splatAttr.getSplatValue();
+ auto splatElementAttr = splatAttr.getSplatValue<Attribute>();
auto splatValue = rewriter.create<arith::ConstantOp>(
constantOp.getLoc(), splatElementAttr.getType(), splatElementAttr);
auto resultType = IREE::Stream::ResourceType::get(constantOp.getContext());
diff --git a/iree/compiler/Dialect/Util/IR/UtilAttrs.cpp b/iree/compiler/Dialect/Util/IR/UtilAttrs.cpp
index f85c30d..bc7a6d8 100644
--- a/iree/compiler/Dialect/Util/IR/UtilAttrs.cpp
+++ b/iree/compiler/Dialect/Util/IR/UtilAttrs.cpp
@@ -496,7 +496,7 @@
auto elementsAttr = baseAttr.cast<DenseElementsAttr>();
if (elementsAttr.isSplat()) {
// Fast-path for splat (no need to convert the value a bunch).
- return serializeSplatValue(elementsAttr.getSplatValue(),
+ return serializeSplatValue(elementsAttr.getSplatValue<Attribute>(),
elementsAttr.getNumElements(), endian, os);
}
diff --git a/iree/compiler/Dialect/VM/IR/VMOpFolders.cpp b/iree/compiler/Dialect/VM/IR/VMOpFolders.cpp
index b930271..64c4791 100644
--- a/iree/compiler/Dialect/VM/IR/VMOpFolders.cpp
+++ b/iree/compiler/Dialect/VM/IR/VMOpFolders.cpp
@@ -542,8 +542,8 @@
if (auto operand = operands[0].dyn_cast_or_null<AttrElementT>()) {
return AttrElementT::get(operand.getType(), calculate(operand.getValue()));
} else if (auto operand = operands[0].dyn_cast_or_null<SplatElementsAttr>()) {
- auto elementResult =
- constFoldUnaryOp<AttrElementT>({operand.getSplatValue()}, calculate);
+ auto elementResult = constFoldUnaryOp<AttrElementT>(
+ {operand.getSplatValue<Attribute>()}, calculate);
if (!elementResult) return {};
return DenseElementsAttr::get(operand.getType(), elementResult);
} else if (auto operand = operands[0].dyn_cast_or_null<ElementsAttr>()) {
@@ -565,7 +565,7 @@
return FloatAttr::get(operand.getType(), calculate(operand.getValue()));
} else if (auto operand = operands[0].dyn_cast_or_null<SplatElementsAttr>()) {
auto elementResult =
- constFoldFloatUnaryOp({operand.getSplatValue()}, calculate);
+ constFoldFloatUnaryOp({operand.getSplatValue<Attribute>()}, calculate);
if (!elementResult) return {};
return DenseElementsAttr::get(operand.getType(), elementResult);
} else if (auto operand = operands[0].dyn_cast_or_null<ElementsAttr>()) {
@@ -597,7 +597,8 @@
auto rhs = operands[1].dyn_cast_or_null<SplatElementsAttr>();
if (!rhs || lhs.getType() != rhs.getType()) return {};
auto elementResult = constFoldBinaryOp<AttrElementT>(
- {lhs.getSplatValue(), rhs.getSplatValue()}, calculate);
+ {lhs.getSplatValue<Attribute>(), rhs.getSplatValue<Attribute>()},
+ calculate);
if (!elementResult) return {};
return DenseElementsAttr::get(lhs.getType(), elementResult);
} else if (auto lhs = operands[0].dyn_cast_or_null<ElementsAttr>()) {
@@ -643,7 +644,9 @@
return {};
}
auto elementResult = constFoldTernaryOp<AttrElementT>(
- {a.getSplatValue(), b.getSplatValue(), c.getSplatValue()}, calculate);
+ {a.getSplatValue<Attribute>(), b.getSplatValue<Attribute>(),
+ c.getSplatValue<Attribute>()},
+ calculate);
if (!elementResult) return {};
return DenseElementsAttr::get(a.getType(), elementResult);
} else if (auto a = operands[0].dyn_cast_or_null<ElementsAttr>()) {
diff --git a/iree/compiler/Dialect/VM/IR/VMOps.cpp b/iree/compiler/Dialect/VM/IR/VMOps.cpp
index 86b7eb4..35213d0 100644
--- a/iree/compiler/Dialect/VM/IR/VMOps.cpp
+++ b/iree/compiler/Dialect/VM/IR/VMOps.cpp
@@ -541,7 +541,8 @@
dims = v.getNumElements();
ShapedType adjustedType = VectorType::get({dims}, integerType);
if (auto elements = v.dyn_cast<SplatElementsAttr>()) {
- return SplatElementsAttr::get(adjustedType, elements.getSplatValue());
+ return SplatElementsAttr::get(adjustedType,
+ elements.getSplatValue<Attribute>());
} else {
return DenseElementsAttr::get(
adjustedType, llvm::to_vector<4>(v.getValues<Attribute>()));
@@ -577,7 +578,8 @@
dims = v.getNumElements();
ShapedType adjustedType = VectorType::get({dims}, floatType);
if (auto elements = v.dyn_cast<SplatElementsAttr>()) {
- return SplatElementsAttr::get(adjustedType, elements.getSplatValue());
+ return SplatElementsAttr::get(adjustedType,
+ elements.getSplatValue<Attribute>());
} else {
return DenseElementsAttr::get(
adjustedType, llvm::to_vector<4>(v.getValues<Attribute>()));
diff --git a/iree/compiler/InputConversion/MHLO/ConvertMHLOToLinalgExt.cpp b/iree/compiler/InputConversion/MHLO/ConvertMHLOToLinalgExt.cpp
index 756b9bb..48113ad 100644
--- a/iree/compiler/InputConversion/MHLO/ConvertMHLOToLinalgExt.cpp
+++ b/iree/compiler/InputConversion/MHLO/ConvertMHLOToLinalgExt.cpp
@@ -395,8 +395,7 @@
if (!operandType || !operandType.hasStaticShape()) {
return failure();
}
- int fftLength =
- op.fft_length().getSplatValue().cast<IntegerAttr>().getInt();
+ int fftLength = op.fft_length().getSplatValue<IntegerAttr>().getInt();
if (fftLength & (fftLength - 1)) {
return rewriter.notifyMatchFailure(
op, "expected FFT length to be a power of two");
diff --git a/iree/compiler/InputConversion/MHLO/LegalizeInputTypes.cpp b/iree/compiler/InputConversion/MHLO/LegalizeInputTypes.cpp
index d765b77..5cf8e42 100644
--- a/iree/compiler/InputConversion/MHLO/LegalizeInputTypes.cpp
+++ b/iree/compiler/InputConversion/MHLO/LegalizeInputTypes.cpp
@@ -56,7 +56,7 @@
} else if (auto attr = value.dyn_cast<SplatElementsAttr>()) {
return SplatElementsAttr::get(
newType.cast<ShapedType>(),
- convertAttribute(loc, attr.getSplatValue(), typeConverter));
+ convertAttribute(loc, attr.getSplatValue<Attribute>(), typeConverter));
} else if (auto attr = value.dyn_cast<DenseIntElementsAttr>()) {
auto newElementType = newType.cast<ShapedType>().getElementType();
auto newElementBitWidth = newElementType.getIntOrFloatBitWidth();
diff --git a/iree/compiler/InputConversion/MHLO/MHLOToLinalgOnTensors.cpp b/iree/compiler/InputConversion/MHLO/MHLOToLinalgOnTensors.cpp
index b1289fa..6a61233 100644
--- a/iree/compiler/InputConversion/MHLO/MHLOToLinalgOnTensors.cpp
+++ b/iree/compiler/InputConversion/MHLO/MHLOToLinalgOnTensors.cpp
@@ -183,7 +183,7 @@
int rank = inputType.getRank();
int n = inputType.getDimSize(rank - 1);
int fftLength =
- op.fft_length().getSplatValue().cast<IntegerAttr>().getInt() / 2 + 1;
+ op.fft_length().getSplatValue<IntegerAttr>().getInt() / 2 + 1;
Location loc = op.getLoc();
auto matrixType =
diff --git a/iree/compiler/InputConversion/MHLO/MHLOToMHLOPreprocessing.cpp b/iree/compiler/InputConversion/MHLO/MHLOToMHLOPreprocessing.cpp
index abf39f2..bb1d5b9 100644
--- a/iree/compiler/InputConversion/MHLO/MHLOToMHLOPreprocessing.cpp
+++ b/iree/compiler/InputConversion/MHLO/MHLOToMHLOPreprocessing.cpp
@@ -145,8 +145,8 @@
llvm::enumerate(op.dimension_numbers().getInputSpatialDimensions())) {
unsigned idx = iter.index();
unsigned dim = iter.value();
- paddingLow[dim] = op.paddingAttr().getValue<int64_t>({idx, 0});
- paddingHigh[dim] = op.paddingAttr().getValue<int64_t>({idx, 1});
+ paddingLow[dim] = op.paddingAttr().getValues<int64_t>()[{idx, 0}];
+ paddingHigh[dim] = op.paddingAttr().getValues<int64_t>()[{idx, 1}];
}
for (unsigned i = 0; i < rank; ++i) {
// mhlo.pad doesn't support dynamic shape.
@@ -402,8 +402,8 @@
SmallVector<int64_t, 4> paddingLow, paddingHigh, interiorPadding, shape;
for (unsigned i = 0; i < rank; ++i) {
interiorPadding.push_back(0);
- paddingLow.push_back(op.paddingAttr().getValue<int64_t>({i, 0}));
- paddingHigh.push_back(op.paddingAttr().getValue<int64_t>({i, 1}));
+ paddingLow.push_back(op.paddingAttr().getValues<int64_t>()[{i, 0}]);
+ paddingHigh.push_back(op.paddingAttr().getValues<int64_t>()[{i, 1}]);
int size = inputShape[i];
shape.push_back(size + paddingLow.back() + paddingHigh.back());
}
diff --git a/third_party/llvm-project b/third_party/llvm-project
index ca47447..6706a47 160000
--- a/third_party/llvm-project
+++ b/third_party/llvm-project
@@ -1 +1 @@
-Subproject commit ca47447952f1f8b0de11aac75b45f83f88579b80
+Subproject commit 6706a4720ffc229eb0b606728b26e988a98c6262
