[flang] update fir.coordinate_of to carry the fields #127231
Merged
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
jeanPerier
requested review from
razvanlupusoru,
rscottmanley,
clementval,
tblah and
vzakhari
|
Thanks for the patch @jeanPerier! The intent to remove such operation looks good to me. |
|
Nice work! |
clementval
approved these changes
Feb 23, 2025
tblah
reviewed
Feb 24, 2025
jeanPerier
force-pushed
the
users/jeanPerier/no-field-index
branch
from
e6a487e to
538552d
Compare
jeanPerier
changed the title
llvmbot
added
flang
|
@llvm/pr-subscribers-flang-codegen @llvm/pr-subscribers-flang-fir-hlfir Author: None (jeanPerier) ChangesThis patch updates fir.coordinate_op to carry the field index as attributes instead of relying on getting it from the fir.field_index operations defining its operands. The rational is that FIR currently has a few operations that require DAGs to be preserved in order to be able to do code generation. This is the case of fir.coordinate_op, which requires its fir.field operand producer to be visible. This make IR transformation harder/brittle, so I want to update FIR to get rid if this. Codegen/printer/parser of fir.coordinate_of and many tests need to be updated after this change. After this patch similar changes should be done to make FIR more robust:
Patch is 424.24 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/127231.diff 68 Files Affected:
diff --git a/flang/include/flang/Optimizer/Dialect/FIROps.h b/flang/include/flang/Optimizer/Dialect/FIROps.h
index a21f8bbe17685..ed301016ad01c 100644
--- a/flang/include/flang/Optimizer/Dialect/FIROps.h
+++ b/flang/include/flang/Optimizer/Dialect/FIROps.h
@@ -50,9 +50,95 @@ struct DebuggingResource
mlir::StringRef getName() final { return "DebuggingResource"; }
};
+class CoordinateIndicesAdaptor;
+using IntOrValue = llvm::PointerUnion<mlir::IntegerAttr, mlir::Value>;
+
} // namespace fir
#define GET_OP_CLASSES
#include "flang/Optimizer/Dialect/FIROps.h.inc"
+namespace fir {
+class CoordinateIndicesAdaptor {
+public:
+ using value_type = IntOrValue;
+
+ CoordinateIndicesAdaptor(mlir::DenseI32ArrayAttr fieldIndices,
+ mlir::ValueRange values)
+ : fieldIndices(fieldIndices), values(values) {}
+
+ value_type operator[](size_t index) const {
+ assert(index < size() && "index out of bounds");
+ return *std::next(begin(), index);
+ }
+
+ size_t size() const {
+ return fieldIndices ? fieldIndices.size() : values.size();
+ }
+
+ bool empty() const {
+ return values.empty() && (!fieldIndices || fieldIndices.empty());
+ }
+
+ class iterator
+ : public llvm::iterator_facade_base<iterator, std::forward_iterator_tag,
+ value_type, std::ptrdiff_t,
+ value_type *, value_type> {
+ public:
+ iterator(const CoordinateIndicesAdaptor *base,
+ std::optional<llvm::ArrayRef<int32_t>::iterator> fieldIter,
+ llvm::detail::IterOfRange<const mlir::ValueRange> valuesIter)
+ : base(base), fieldIter(fieldIter), valuesIter(valuesIter) {}
+
+ value_type operator*() const {
+ if (fieldIter && **fieldIter != fir::CoordinateOp::kDynamicIndex) {
+ return mlir::IntegerAttr::get(base->fieldIndices.getElementType(),
+ **fieldIter);
+ }
+ return *valuesIter;
+ }
+
+ iterator &operator++() {
+ if (fieldIter) {
+ if (**fieldIter == fir::CoordinateOp::kDynamicIndex)
+ valuesIter++;
+ (*fieldIter)++;
+ } else {
+ valuesIter++;
+ }
+ return *this;
+ }
+
+ bool operator==(const iterator &rhs) const {
+ return base == rhs.base && fieldIter == rhs.fieldIter &&
+ valuesIter == rhs.valuesIter;
+ }
+
+ private:
+ const CoordinateIndicesAdaptor *base;
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ llvm::detail::IterOfRange<const mlir::ValueRange> valuesIter;
+ };
+
+ iterator begin() const {
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ if (fieldIndices)
+ fieldIter = fieldIndices.asArrayRef().begin();
+ return iterator(this, fieldIter, values.begin());
+ }
+
+ iterator end() const {
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ if (fieldIndices)
+ fieldIter = fieldIndices.asArrayRef().end();
+ return iterator(this, fieldIter, values.end());
+ }
+
+private:
+ mlir::DenseI32ArrayAttr fieldIndices;
+ mlir::ValueRange values;
+};
+
+} // namespace fir
+
#endif // FORTRAN_OPTIMIZER_DIALECT_FIROPS_H
diff --git a/flang/include/flang/Optimizer/Dialect/FIROps.td b/flang/include/flang/Optimizer/Dialect/FIROps.td
index 8dbc9df9f553d..c83c57186b46d 100644
--- a/flang/include/flang/Optimizer/Dialect/FIROps.td
+++ b/flang/include/flang/Optimizer/Dialect/FIROps.td
@@ -1748,10 +1748,16 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
Unlike LLVM's GEP instruction, one cannot stride over the outermost
reference; therefore, the leading 0 index must be omitted.
+ This operation can be used to index derived type fields, in which case
+ the operand is the name of the index field.
+
```
%i = ... : index
%h = ... : !fir.heap<!fir.array<100 x f32>>
%p = fir.coordinate_of %h, %i : (!fir.heap<!fir.array<100 x f32>>, index) -> !fir.ref<f32>
+
+ %d = ... : !fir.ref<!fir.type<t{field1:i32, field2:f32}>>
+ %f = fir.coordinate_of %d, field2 : (!fir.ref<!fir.type<t{field1:i32, field2:f32}>>) -> !fir.ref<f32>
```
In the example, `%p` will be a pointer to the `%i`-th f32 value in the
@@ -1761,7 +1767,8 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
let arguments = (ins
AnyRefOrBox:$ref,
Variadic<AnyCoordinateType>:$coor,
- TypeAttr:$baseType
+ TypeAttr:$baseType,
+ OptionalAttr<DenseI32ArrayAttr>:$field_indices
);
let results = (outs RefOrLLVMPtr);
@@ -1771,10 +1778,14 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
let builders = [
OpBuilder<(ins "mlir::Type":$resultType,
- "mlir::Value":$ref, "mlir::ValueRange":$coor),
- [{ return build($_builder, $_state, resultType, ref, coor,
- mlir::TypeAttr::get(ref.getType())); }]>,
+ "mlir::Value":$ref, "mlir::ValueRange":$coor)>,
+ OpBuilder<(ins "mlir::Type":$resultType,
+ "mlir::Value":$ref, "llvm::ArrayRef<fir::IntOrValue>":$coor)>
];
+ let extraClassDeclaration = [{
+ constexpr static int32_t kDynamicIndex = std::numeric_limits<int32_t>::min();
+ CoordinateIndicesAdaptor getIndices();
+ }];
}
def fir_ExtractValueOp : fir_OneResultOp<"extract_value", [NoMemoryEffect]> {
diff --git a/flang/lib/Lower/OpenMP/Utils.cpp b/flang/lib/Lower/OpenMP/Utils.cpp
index fa1975dac789b..48bcf492fd368 100644
--- a/flang/lib/Lower/OpenMP/Utils.cpp
+++ b/flang/lib/Lower/OpenMP/Utils.cpp
@@ -354,14 +354,12 @@ mlir::Value createParentSymAndGenIntermediateMaps(
// type.
if (fir::RecordType recordType = mlir::dyn_cast<fir::RecordType>(
fir::unwrapPassByRefType(curValue.getType()))) {
- mlir::Value idxConst = firOpBuilder.createIntegerConstant(
- clauseLocation, firOpBuilder.getIndexType(),
- indices[currentIndicesIdx]);
- mlir::Type memberTy =
- recordType.getTypeList().at(indices[currentIndicesIdx]).second;
+ fir::IntOrValue idxConst = mlir::IntegerAttr::get(
+ firOpBuilder.getI32Type(), indices[currentIndicesIdx]);
+ mlir::Type memberTy = recordType.getType(indices[currentIndicesIdx]);
curValue = firOpBuilder.create<fir::CoordinateOp>(
clauseLocation, firOpBuilder.getRefType(memberTy), curValue,
- idxConst);
+ llvm::SmallVector<fir::IntOrValue, 1>{idxConst});
// Skip mapping and the subsequent load if we're the final member or not
// a type with a descriptor such as a pointer/allocatable. If we're a
diff --git a/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp b/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
index 26f4aee21d8bd..82b11ad7db32a 100644
--- a/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
+++ b/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
@@ -348,8 +348,9 @@ class BoxedProcedurePass
rewriter.setInsertionPoint(coor);
auto toTy = typeConverter.convertType(ty);
auto toBaseTy = typeConverter.convertType(baseTy);
- rewriter.replaceOpWithNewOp<CoordinateOp>(coor, toTy, coor.getRef(),
- coor.getCoor(), toBaseTy);
+ rewriter.replaceOpWithNewOp<CoordinateOp>(
+ coor, toTy, coor.getRef(), coor.getCoor(), toBaseTy,
+ coor.getFieldIndicesAttr());
opIsValid = false;
}
} else if (auto index = mlir::dyn_cast<FieldIndexOp>(op)) {
diff --git a/flang/lib/Optimizer/CodeGen/CodeGen.cpp b/flang/lib/Optimizer/CodeGen/CodeGen.cpp
index aaefe675730e1..a2743edd7844a 100644
--- a/flang/lib/Optimizer/CodeGen/CodeGen.cpp
+++ b/flang/lib/Optimizer/CodeGen/CodeGen.cpp
@@ -2653,57 +2653,78 @@ struct CoordinateOpConversion
return mlir::isa<fir::SequenceType, fir::RecordType, mlir::TupleType>(type);
}
- /// Check whether this form of `!fir.coordinate_of` is supported. These
- /// additional checks are required, because we are not yet able to convert
- /// all valid forms of `!fir.coordinate_of`.
- /// TODO: Either implement the unsupported cases or extend the verifier
- /// in FIROps.cpp instead.
- static bool supportedCoordinate(mlir::Type type, mlir::ValueRange coors) {
- const std::size_t numOfCoors = coors.size();
- std::size_t i = 0;
- bool subEle = false;
- bool ptrEle = false;
- for (; i < numOfCoors; ++i) {
- mlir::Value nxtOpnd = coors[i];
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(type)) {
- subEle = true;
- i += arrTy.getDimension() - 1;
- type = arrTy.getEleTy();
- } else if (auto recTy = mlir::dyn_cast<fir::RecordType>(type)) {
- subEle = true;
- type = recTy.getType(getFieldNumber(recTy, nxtOpnd));
- } else if (auto tupTy = mlir::dyn_cast<mlir::TupleType>(type)) {
- subEle = true;
- type = tupTy.getType(getConstantIntValue(nxtOpnd));
- } else {
- ptrEle = true;
- }
- }
- if (ptrEle)
- return (!subEle) && (numOfCoors == 1);
- return subEle && (i >= numOfCoors);
- }
+ // Helper structure to analyze the CoordinateOp path and decide if and how
+ // the GEP should be generated for it.
+ struct ShapeAnalysis {
+ bool hasKnownShape;
+ bool columnIsDeferred;
+ };
/// Walk the abstract memory layout and determine if the path traverses any
/// array types with unknown shape. Return true iff all the array types have a
/// constant shape along the path.
- static bool arraysHaveKnownShape(mlir::Type type, mlir::ValueRange coors) {
- for (std::size_t i = 0, sz = coors.size(); i < sz; ++i) {
- mlir::Value nxtOpnd = coors[i];
+ /// TODO: move the verification logic into the verifier.
+ static std::optional<ShapeAnalysis>
+ arraysHaveKnownShape(mlir::Type type, fir::CoordinateOp coor) {
+ fir::CoordinateIndicesAdaptor indices = coor.getIndices();
+ auto begin = indices.begin();
+ bool hasKnownShape = true;
+ bool columnIsDeferred = false;
+ for (auto it = begin, end = indices.end(); it != end;) {
if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(type)) {
- if (fir::sequenceWithNonConstantShape(arrTy))
- return false;
- i += arrTy.getDimension() - 1;
+ bool addressingStart = (it == begin);
+ unsigned arrayDim = arrTy.getDimension();
+ for (auto dimExtent : llvm::enumerate(arrTy.getShape())) {
+ if (dimExtent.value() == fir::SequenceType::getUnknownExtent()) {
+ hasKnownShape = false;
+ if (addressingStart && dimExtent.index() + 1 == arrayDim) {
+ // If this point was reached, the raws of the first array have
+ // constant extents.
+ columnIsDeferred = true;
+ } else {
+ // One of the array dimension that is not the column of the first
+ // array has dynamic extent. It will not possible to do
+ // code generation for the CoordinateOp if the base is not a
+ // fir.box containing the value of that extent.
+ return ShapeAnalysis{false, false};
+ }
+ }
+ // There may be less operands than the array size if the
+ // fir.coordinate_of result is not an element but a sub-array.
+ if (it != end)
+ ++it;
+ }
type = arrTy.getEleTy();
- } else if (auto strTy = mlir::dyn_cast<fir::RecordType>(type)) {
- type = strTy.getType(getFieldNumber(strTy, nxtOpnd));
+ continue;
+ }
+ if (auto strTy = mlir::dyn_cast<fir::RecordType>(type)) {
+ auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(*it);
+ if (!intAttr) {
+ mlir::emitError(coor.getLoc(),
+ "expected field name in fir.coordinate_of");
+ return std::nullopt;
+ }
+ type = strTy.getType(intAttr.getInt());
} else if (auto strTy = mlir::dyn_cast<mlir::TupleType>(type)) {
- type = strTy.getType(getConstantIntValue(nxtOpnd));
- } else {
- return true;
+ auto value = llvm::dyn_cast<mlir::Value>(*it);
+ if (!value) {
+ mlir::emitError(
+ coor.getLoc(),
+ "expected constant value to address tuple in fir.coordinate_of");
+ return std::nullopt;
+ }
+ type = strTy.getType(getConstantIntValue(value));
+ } else if (auto charType = mlir::dyn_cast<fir::CharacterType>(type)) {
+ // Addressing character in string. Fortran strings degenerate to arrays
+ // in LLVM, so they are handled like arrays of characters here.
+ if (charType.getLen() == fir::CharacterType::unknownLen())
+ return ShapeAnalysis{false, true};
+ type = fir::CharacterType::getSingleton(charType.getContext(),
+ charType.getFKind());
}
+ ++it;
}
- return true;
+ return ShapeAnalysis{hasKnownShape, columnIsDeferred};
}
private:
@@ -2754,9 +2775,11 @@ struct CoordinateOpConversion
mlir::LLVM::IntegerOverflowFlags nsw =
mlir::LLVM::IntegerOverflowFlags::nsw;
- for (unsigned i = 1, last = operands.size(); i < last; ++i) {
+ int nextIndexValue = 1;
+ fir::CoordinateIndicesAdaptor indices = coor.getIndices();
+ for (auto it = indices.begin(), end = indices.end(); it != end;) {
if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- if (i != 1)
+ if (it != indices.begin())
TODO(loc, "fir.array nested inside other array and/or derived type");
// Applies byte strides from the box. Ignore lower bound from box
// since fir.coordinate_of indexes are zero based. Lowering takes care
@@ -2764,26 +2787,31 @@ struct CoordinateOpConversion
// types and non contiguous arrays.
auto idxTy = lowerTy().indexType();
mlir::Value off = genConstantIndex(loc, idxTy, rewriter, 0);
- for (unsigned index = i, lastIndex = i + arrTy.getDimension();
- index < lastIndex; ++index) {
- mlir::Value stride = getStrideFromBox(loc, boxTyPair, operands[0],
- index - i, rewriter);
+ unsigned arrayDim = arrTy.getDimension();
+ for (unsigned dim = 0; dim < arrayDim && it != end; ++dim, ++it) {
+ mlir::Value stride =
+ getStrideFromBox(loc, boxTyPair, operands[0], dim, rewriter);
auto sc = rewriter.create<mlir::LLVM::MulOp>(
- loc, idxTy, operands[index], stride, nsw);
+ loc, idxTy, operands[nextIndexValue + dim], stride, nsw);
off = rewriter.create<mlir::LLVM::AddOp>(loc, idxTy, sc, off, nsw);
}
+ nextIndexValue += arrayDim;
resultAddr = rewriter.create<mlir::LLVM::GEPOp>(
loc, llvmPtrTy, byteTy, resultAddr,
llvm::ArrayRef<mlir::LLVM::GEPArg>{off});
- i += arrTy.getDimension() - 1;
cpnTy = arrTy.getEleTy();
} else if (auto recTy = mlir::dyn_cast<fir::RecordType>(cpnTy)) {
- mlir::Value nxtOpnd = operands[i];
- cpnTy = recTy.getType(getFieldNumber(recTy, nxtOpnd));
+ auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(*it);
+ if (!intAttr)
+ return mlir::emitError(loc,
+ "expected field name in fir.coordinate_of");
+ int fieldIndex = intAttr.getInt();
+ ++it;
+ cpnTy = recTy.getType(fieldIndex);
auto llvmRecTy = lowerTy().convertType(recTy);
resultAddr = rewriter.create<mlir::LLVM::GEPOp>(
loc, llvmPtrTy, llvmRecTy, resultAddr,
- llvm::ArrayRef<mlir::LLVM::GEPArg>{0, nxtOpnd});
+ llvm::ArrayRef<mlir::LLVM::GEPArg>{0, fieldIndex});
} else {
fir::emitFatalError(loc, "unexpected type in coordinate_of");
}
@@ -2801,92 +2829,71 @@ struct CoordinateOpConversion
// Component Type
mlir::Type cpnTy = fir::dyn_cast_ptrOrBoxEleTy(baseObjectTy);
- bool hasSubdimension = hasSubDimensions(cpnTy);
- bool columnIsDeferred = !hasSubdimension;
-
- if (!supportedCoordinate(cpnTy, operands.drop_front(1)))
- TODO(loc, "unsupported combination of coordinate operands");
-
- const bool hasKnownShape =
- arraysHaveKnownShape(cpnTy, operands.drop_front(1));
-
- // If only the column is `?`, then we can simply place the column value in
- // the 0-th GEP position.
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- if (!hasKnownShape) {
- const unsigned sz = arrTy.getDimension();
- if (arraysHaveKnownShape(arrTy.getEleTy(),
- operands.drop_front(1 + sz))) {
- fir::SequenceType::ShapeRef shape = arrTy.getShape();
- bool allConst = true;
- for (unsigned i = 0; i < sz - 1; ++i) {
- if (shape[i] < 0) {
- allConst = false;
- break;
- }
- }
- if (allConst)
- columnIsDeferred = true;
- }
- }
- }
+
+ const std::optional<ShapeAnalysis> shapeAnalysis =
+ arraysHaveKnownShape(cpnTy, coor);
+ if (!shapeAnalysis)
+ return mlir::failure();
if (fir::hasDynamicSize(fir::unwrapSequenceType(cpnTy)))
return mlir::emitError(
loc, "fir.coordinate_of with a dynamic element size is unsupported");
- if (hasKnownShape || columnIsDeferred) {
+ if (shapeAnalysis->hasKnownShape || shapeAnalysis->columnIsDeferred) {
llvm::SmallVector<mlir::LLVM::GEPArg> offs;
- if (hasKnownShape && hasSubdimension) {
+ if (shapeAnalysis->hasKnownShape) {
offs.push_back(0);
}
+ // Else, only the column is `?` and we can simply place the column value
+ // in the 0-th GEP position.
+
std::optional<int> dims;
llvm::SmallVector<mlir::Value> arrIdx;
- for (std::size_t i = 1, sz = operands.size(); i < sz; ++i) {
- mlir::Value nxtOpnd = operands[i];
-
- if (!cpnTy)
- return mlir::emitError(loc, "invalid coordinate/check failed");
-
- // check if the i-th coordinate relates to an array
- if (dims) {
- arrIdx.push_back(nxtOpnd);
- int dimsLeft = *dims;
- if (dimsLeft > 1) {
- dims = dimsLeft - 1;
- continue;
- }
- cpnTy = mlir::cast<fir::SequenceType>(cpnTy).getElementType();
- // append array range in reverse (FIR arrays are column-major)
- offs.append(arrIdx.rbegin(), arrIdx.rend());
- arrIdx.clear();
- dims.reset();
+ int nextIndexValue = 1;
+ for (auto index : coor.getIndices()) {
+ if (auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(index)) {
+ // Addressing derived type component.
+ auto recordType = llvm::dyn_cast<fir::RecordType>(cpnTy);
+ if (!recordType)
+ return mlir::emitError(
+ loc,
+ "fir.coordinate base type is not consistent with operands");
+ int fieldId = intAttr.getInt();
+ cpnTy = recordType.getType(fieldId);
+ offs.push_back(fieldId);
continue;
}
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- int d = arrTy.getDimension() - 1;
- if (d > 0) {
- dims = d;
- arrIdx.push_back(nxtOpnd);
- continue;
+ // Value index (addressing array, tuple, or complex part).
+ mlir::Value indexValue = operands[nextIndexValue++];
+ if (auto tupTy = mlir::dyn_cast<mlir::TupleType>(cpnTy)) {
+ cpnTy = tupTy.getType(getConstantIntValue(indexValue));
+ offs.push_back(indexValue);
+ } else {
+ if (!dims) {
+ if (auto arrayType = llvm::dyn_cast<fir::SequenceType>(cpnTy)) {
+ // Starting addressing array or array component.
+ dims = arrayType.getDimension();
+ cpnTy = arrayType.getElementType();
+ }
+ }
+ if (dims) {
+ arrIdx.push_back(indexValue);
+ if (--(*dims) == 0) {
+ // Append array range in reverse (FIR arrays are column-major).
+ offs.append(arrIdx.rbegin(), arrIdx.rend());
+ arrIdx.clear();
+ dims.reset();
+ ...
[truncated]
|
|
@llvm/pr-subscribers-flang-openmp Author: None (jeanPerier) ChangesThis patch updates fir.coordinate_op to carry the field index as attributes instead of relying on getting it from the fir.field_index operations defining its operands. The rational is that FIR currently has a few operations that require DAGs to be preserved in order to be able to do code generation. This is the case of fir.coordinate_op, which requires its fir.field operand producer to be visible. This make IR transformation harder/brittle, so I want to update FIR to get rid if this. Codegen/printer/parser of fir.coordinate_of and many tests need to be updated after this change. After this patch similar changes should be done to make FIR more robust:
Patch is 424.24 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/127231.diff 68 Files Affected:
diff --git a/flang/include/flang/Optimizer/Dialect/FIROps.h b/flang/include/flang/Optimizer/Dialect/FIROps.h
index a21f8bbe17685..ed301016ad01c 100644
--- a/flang/include/flang/Optimizer/Dialect/FIROps.h
+++ b/flang/include/flang/Optimizer/Dialect/FIROps.h
@@ -50,9 +50,95 @@ struct DebuggingResource
mlir::StringRef getName() final { return "DebuggingResource"; }
};
+class CoordinateIndicesAdaptor;
+using IntOrValue = llvm::PointerUnion<mlir::IntegerAttr, mlir::Value>;
+
} // namespace fir
#define GET_OP_CLASSES
#include "flang/Optimizer/Dialect/FIROps.h.inc"
+namespace fir {
+class CoordinateIndicesAdaptor {
+public:
+ using value_type = IntOrValue;
+
+ CoordinateIndicesAdaptor(mlir::DenseI32ArrayAttr fieldIndices,
+ mlir::ValueRange values)
+ : fieldIndices(fieldIndices), values(values) {}
+
+ value_type operator[](size_t index) const {
+ assert(index < size() && "index out of bounds");
+ return *std::next(begin(), index);
+ }
+
+ size_t size() const {
+ return fieldIndices ? fieldIndices.size() : values.size();
+ }
+
+ bool empty() const {
+ return values.empty() && (!fieldIndices || fieldIndices.empty());
+ }
+
+ class iterator
+ : public llvm::iterator_facade_base<iterator, std::forward_iterator_tag,
+ value_type, std::ptrdiff_t,
+ value_type *, value_type> {
+ public:
+ iterator(const CoordinateIndicesAdaptor *base,
+ std::optional<llvm::ArrayRef<int32_t>::iterator> fieldIter,
+ llvm::detail::IterOfRange<const mlir::ValueRange> valuesIter)
+ : base(base), fieldIter(fieldIter), valuesIter(valuesIter) {}
+
+ value_type operator*() const {
+ if (fieldIter && **fieldIter != fir::CoordinateOp::kDynamicIndex) {
+ return mlir::IntegerAttr::get(base->fieldIndices.getElementType(),
+ **fieldIter);
+ }
+ return *valuesIter;
+ }
+
+ iterator &operator++() {
+ if (fieldIter) {
+ if (**fieldIter == fir::CoordinateOp::kDynamicIndex)
+ valuesIter++;
+ (*fieldIter)++;
+ } else {
+ valuesIter++;
+ }
+ return *this;
+ }
+
+ bool operator==(const iterator &rhs) const {
+ return base == rhs.base && fieldIter == rhs.fieldIter &&
+ valuesIter == rhs.valuesIter;
+ }
+
+ private:
+ const CoordinateIndicesAdaptor *base;
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ llvm::detail::IterOfRange<const mlir::ValueRange> valuesIter;
+ };
+
+ iterator begin() const {
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ if (fieldIndices)
+ fieldIter = fieldIndices.asArrayRef().begin();
+ return iterator(this, fieldIter, values.begin());
+ }
+
+ iterator end() const {
+ std::optional<llvm::ArrayRef<int32_t>::const_iterator> fieldIter;
+ if (fieldIndices)
+ fieldIter = fieldIndices.asArrayRef().end();
+ return iterator(this, fieldIter, values.end());
+ }
+
+private:
+ mlir::DenseI32ArrayAttr fieldIndices;
+ mlir::ValueRange values;
+};
+
+} // namespace fir
+
#endif // FORTRAN_OPTIMIZER_DIALECT_FIROPS_H
diff --git a/flang/include/flang/Optimizer/Dialect/FIROps.td b/flang/include/flang/Optimizer/Dialect/FIROps.td
index 8dbc9df9f553d..c83c57186b46d 100644
--- a/flang/include/flang/Optimizer/Dialect/FIROps.td
+++ b/flang/include/flang/Optimizer/Dialect/FIROps.td
@@ -1748,10 +1748,16 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
Unlike LLVM's GEP instruction, one cannot stride over the outermost
reference; therefore, the leading 0 index must be omitted.
+ This operation can be used to index derived type fields, in which case
+ the operand is the name of the index field.
+
```
%i = ... : index
%h = ... : !fir.heap<!fir.array<100 x f32>>
%p = fir.coordinate_of %h, %i : (!fir.heap<!fir.array<100 x f32>>, index) -> !fir.ref<f32>
+
+ %d = ... : !fir.ref<!fir.type<t{field1:i32, field2:f32}>>
+ %f = fir.coordinate_of %d, field2 : (!fir.ref<!fir.type<t{field1:i32, field2:f32}>>) -> !fir.ref<f32>
```
In the example, `%p` will be a pointer to the `%i`-th f32 value in the
@@ -1761,7 +1767,8 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
let arguments = (ins
AnyRefOrBox:$ref,
Variadic<AnyCoordinateType>:$coor,
- TypeAttr:$baseType
+ TypeAttr:$baseType,
+ OptionalAttr<DenseI32ArrayAttr>:$field_indices
);
let results = (outs RefOrLLVMPtr);
@@ -1771,10 +1778,14 @@ def fir_CoordinateOp : fir_Op<"coordinate_of", [NoMemoryEffect]> {
let builders = [
OpBuilder<(ins "mlir::Type":$resultType,
- "mlir::Value":$ref, "mlir::ValueRange":$coor),
- [{ return build($_builder, $_state, resultType, ref, coor,
- mlir::TypeAttr::get(ref.getType())); }]>,
+ "mlir::Value":$ref, "mlir::ValueRange":$coor)>,
+ OpBuilder<(ins "mlir::Type":$resultType,
+ "mlir::Value":$ref, "llvm::ArrayRef<fir::IntOrValue>":$coor)>
];
+ let extraClassDeclaration = [{
+ constexpr static int32_t kDynamicIndex = std::numeric_limits<int32_t>::min();
+ CoordinateIndicesAdaptor getIndices();
+ }];
}
def fir_ExtractValueOp : fir_OneResultOp<"extract_value", [NoMemoryEffect]> {
diff --git a/flang/lib/Lower/OpenMP/Utils.cpp b/flang/lib/Lower/OpenMP/Utils.cpp
index fa1975dac789b..48bcf492fd368 100644
--- a/flang/lib/Lower/OpenMP/Utils.cpp
+++ b/flang/lib/Lower/OpenMP/Utils.cpp
@@ -354,14 +354,12 @@ mlir::Value createParentSymAndGenIntermediateMaps(
// type.
if (fir::RecordType recordType = mlir::dyn_cast<fir::RecordType>(
fir::unwrapPassByRefType(curValue.getType()))) {
- mlir::Value idxConst = firOpBuilder.createIntegerConstant(
- clauseLocation, firOpBuilder.getIndexType(),
- indices[currentIndicesIdx]);
- mlir::Type memberTy =
- recordType.getTypeList().at(indices[currentIndicesIdx]).second;
+ fir::IntOrValue idxConst = mlir::IntegerAttr::get(
+ firOpBuilder.getI32Type(), indices[currentIndicesIdx]);
+ mlir::Type memberTy = recordType.getType(indices[currentIndicesIdx]);
curValue = firOpBuilder.create<fir::CoordinateOp>(
clauseLocation, firOpBuilder.getRefType(memberTy), curValue,
- idxConst);
+ llvm::SmallVector<fir::IntOrValue, 1>{idxConst});
// Skip mapping and the subsequent load if we're the final member or not
// a type with a descriptor such as a pointer/allocatable. If we're a
diff --git a/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp b/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
index 26f4aee21d8bd..82b11ad7db32a 100644
--- a/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
+++ b/flang/lib/Optimizer/CodeGen/BoxedProcedure.cpp
@@ -348,8 +348,9 @@ class BoxedProcedurePass
rewriter.setInsertionPoint(coor);
auto toTy = typeConverter.convertType(ty);
auto toBaseTy = typeConverter.convertType(baseTy);
- rewriter.replaceOpWithNewOp<CoordinateOp>(coor, toTy, coor.getRef(),
- coor.getCoor(), toBaseTy);
+ rewriter.replaceOpWithNewOp<CoordinateOp>(
+ coor, toTy, coor.getRef(), coor.getCoor(), toBaseTy,
+ coor.getFieldIndicesAttr());
opIsValid = false;
}
} else if (auto index = mlir::dyn_cast<FieldIndexOp>(op)) {
diff --git a/flang/lib/Optimizer/CodeGen/CodeGen.cpp b/flang/lib/Optimizer/CodeGen/CodeGen.cpp
index aaefe675730e1..a2743edd7844a 100644
--- a/flang/lib/Optimizer/CodeGen/CodeGen.cpp
+++ b/flang/lib/Optimizer/CodeGen/CodeGen.cpp
@@ -2653,57 +2653,78 @@ struct CoordinateOpConversion
return mlir::isa<fir::SequenceType, fir::RecordType, mlir::TupleType>(type);
}
- /// Check whether this form of `!fir.coordinate_of` is supported. These
- /// additional checks are required, because we are not yet able to convert
- /// all valid forms of `!fir.coordinate_of`.
- /// TODO: Either implement the unsupported cases or extend the verifier
- /// in FIROps.cpp instead.
- static bool supportedCoordinate(mlir::Type type, mlir::ValueRange coors) {
- const std::size_t numOfCoors = coors.size();
- std::size_t i = 0;
- bool subEle = false;
- bool ptrEle = false;
- for (; i < numOfCoors; ++i) {
- mlir::Value nxtOpnd = coors[i];
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(type)) {
- subEle = true;
- i += arrTy.getDimension() - 1;
- type = arrTy.getEleTy();
- } else if (auto recTy = mlir::dyn_cast<fir::RecordType>(type)) {
- subEle = true;
- type = recTy.getType(getFieldNumber(recTy, nxtOpnd));
- } else if (auto tupTy = mlir::dyn_cast<mlir::TupleType>(type)) {
- subEle = true;
- type = tupTy.getType(getConstantIntValue(nxtOpnd));
- } else {
- ptrEle = true;
- }
- }
- if (ptrEle)
- return (!subEle) && (numOfCoors == 1);
- return subEle && (i >= numOfCoors);
- }
+ // Helper structure to analyze the CoordinateOp path and decide if and how
+ // the GEP should be generated for it.
+ struct ShapeAnalysis {
+ bool hasKnownShape;
+ bool columnIsDeferred;
+ };
/// Walk the abstract memory layout and determine if the path traverses any
/// array types with unknown shape. Return true iff all the array types have a
/// constant shape along the path.
- static bool arraysHaveKnownShape(mlir::Type type, mlir::ValueRange coors) {
- for (std::size_t i = 0, sz = coors.size(); i < sz; ++i) {
- mlir::Value nxtOpnd = coors[i];
+ /// TODO: move the verification logic into the verifier.
+ static std::optional<ShapeAnalysis>
+ arraysHaveKnownShape(mlir::Type type, fir::CoordinateOp coor) {
+ fir::CoordinateIndicesAdaptor indices = coor.getIndices();
+ auto begin = indices.begin();
+ bool hasKnownShape = true;
+ bool columnIsDeferred = false;
+ for (auto it = begin, end = indices.end(); it != end;) {
if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(type)) {
- if (fir::sequenceWithNonConstantShape(arrTy))
- return false;
- i += arrTy.getDimension() - 1;
+ bool addressingStart = (it == begin);
+ unsigned arrayDim = arrTy.getDimension();
+ for (auto dimExtent : llvm::enumerate(arrTy.getShape())) {
+ if (dimExtent.value() == fir::SequenceType::getUnknownExtent()) {
+ hasKnownShape = false;
+ if (addressingStart && dimExtent.index() + 1 == arrayDim) {
+ // If this point was reached, the raws of the first array have
+ // constant extents.
+ columnIsDeferred = true;
+ } else {
+ // One of the array dimension that is not the column of the first
+ // array has dynamic extent. It will not possible to do
+ // code generation for the CoordinateOp if the base is not a
+ // fir.box containing the value of that extent.
+ return ShapeAnalysis{false, false};
+ }
+ }
+ // There may be less operands than the array size if the
+ // fir.coordinate_of result is not an element but a sub-array.
+ if (it != end)
+ ++it;
+ }
type = arrTy.getEleTy();
- } else if (auto strTy = mlir::dyn_cast<fir::RecordType>(type)) {
- type = strTy.getType(getFieldNumber(strTy, nxtOpnd));
+ continue;
+ }
+ if (auto strTy = mlir::dyn_cast<fir::RecordType>(type)) {
+ auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(*it);
+ if (!intAttr) {
+ mlir::emitError(coor.getLoc(),
+ "expected field name in fir.coordinate_of");
+ return std::nullopt;
+ }
+ type = strTy.getType(intAttr.getInt());
} else if (auto strTy = mlir::dyn_cast<mlir::TupleType>(type)) {
- type = strTy.getType(getConstantIntValue(nxtOpnd));
- } else {
- return true;
+ auto value = llvm::dyn_cast<mlir::Value>(*it);
+ if (!value) {
+ mlir::emitError(
+ coor.getLoc(),
+ "expected constant value to address tuple in fir.coordinate_of");
+ return std::nullopt;
+ }
+ type = strTy.getType(getConstantIntValue(value));
+ } else if (auto charType = mlir::dyn_cast<fir::CharacterType>(type)) {
+ // Addressing character in string. Fortran strings degenerate to arrays
+ // in LLVM, so they are handled like arrays of characters here.
+ if (charType.getLen() == fir::CharacterType::unknownLen())
+ return ShapeAnalysis{false, true};
+ type = fir::CharacterType::getSingleton(charType.getContext(),
+ charType.getFKind());
}
+ ++it;
}
- return true;
+ return ShapeAnalysis{hasKnownShape, columnIsDeferred};
}
private:
@@ -2754,9 +2775,11 @@ struct CoordinateOpConversion
mlir::LLVM::IntegerOverflowFlags nsw =
mlir::LLVM::IntegerOverflowFlags::nsw;
- for (unsigned i = 1, last = operands.size(); i < last; ++i) {
+ int nextIndexValue = 1;
+ fir::CoordinateIndicesAdaptor indices = coor.getIndices();
+ for (auto it = indices.begin(), end = indices.end(); it != end;) {
if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- if (i != 1)
+ if (it != indices.begin())
TODO(loc, "fir.array nested inside other array and/or derived type");
// Applies byte strides from the box. Ignore lower bound from box
// since fir.coordinate_of indexes are zero based. Lowering takes care
@@ -2764,26 +2787,31 @@ struct CoordinateOpConversion
// types and non contiguous arrays.
auto idxTy = lowerTy().indexType();
mlir::Value off = genConstantIndex(loc, idxTy, rewriter, 0);
- for (unsigned index = i, lastIndex = i + arrTy.getDimension();
- index < lastIndex; ++index) {
- mlir::Value stride = getStrideFromBox(loc, boxTyPair, operands[0],
- index - i, rewriter);
+ unsigned arrayDim = arrTy.getDimension();
+ for (unsigned dim = 0; dim < arrayDim && it != end; ++dim, ++it) {
+ mlir::Value stride =
+ getStrideFromBox(loc, boxTyPair, operands[0], dim, rewriter);
auto sc = rewriter.create<mlir::LLVM::MulOp>(
- loc, idxTy, operands[index], stride, nsw);
+ loc, idxTy, operands[nextIndexValue + dim], stride, nsw);
off = rewriter.create<mlir::LLVM::AddOp>(loc, idxTy, sc, off, nsw);
}
+ nextIndexValue += arrayDim;
resultAddr = rewriter.create<mlir::LLVM::GEPOp>(
loc, llvmPtrTy, byteTy, resultAddr,
llvm::ArrayRef<mlir::LLVM::GEPArg>{off});
- i += arrTy.getDimension() - 1;
cpnTy = arrTy.getEleTy();
} else if (auto recTy = mlir::dyn_cast<fir::RecordType>(cpnTy)) {
- mlir::Value nxtOpnd = operands[i];
- cpnTy = recTy.getType(getFieldNumber(recTy, nxtOpnd));
+ auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(*it);
+ if (!intAttr)
+ return mlir::emitError(loc,
+ "expected field name in fir.coordinate_of");
+ int fieldIndex = intAttr.getInt();
+ ++it;
+ cpnTy = recTy.getType(fieldIndex);
auto llvmRecTy = lowerTy().convertType(recTy);
resultAddr = rewriter.create<mlir::LLVM::GEPOp>(
loc, llvmPtrTy, llvmRecTy, resultAddr,
- llvm::ArrayRef<mlir::LLVM::GEPArg>{0, nxtOpnd});
+ llvm::ArrayRef<mlir::LLVM::GEPArg>{0, fieldIndex});
} else {
fir::emitFatalError(loc, "unexpected type in coordinate_of");
}
@@ -2801,92 +2829,71 @@ struct CoordinateOpConversion
// Component Type
mlir::Type cpnTy = fir::dyn_cast_ptrOrBoxEleTy(baseObjectTy);
- bool hasSubdimension = hasSubDimensions(cpnTy);
- bool columnIsDeferred = !hasSubdimension;
-
- if (!supportedCoordinate(cpnTy, operands.drop_front(1)))
- TODO(loc, "unsupported combination of coordinate operands");
-
- const bool hasKnownShape =
- arraysHaveKnownShape(cpnTy, operands.drop_front(1));
-
- // If only the column is `?`, then we can simply place the column value in
- // the 0-th GEP position.
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- if (!hasKnownShape) {
- const unsigned sz = arrTy.getDimension();
- if (arraysHaveKnownShape(arrTy.getEleTy(),
- operands.drop_front(1 + sz))) {
- fir::SequenceType::ShapeRef shape = arrTy.getShape();
- bool allConst = true;
- for (unsigned i = 0; i < sz - 1; ++i) {
- if (shape[i] < 0) {
- allConst = false;
- break;
- }
- }
- if (allConst)
- columnIsDeferred = true;
- }
- }
- }
+
+ const std::optional<ShapeAnalysis> shapeAnalysis =
+ arraysHaveKnownShape(cpnTy, coor);
+ if (!shapeAnalysis)
+ return mlir::failure();
if (fir::hasDynamicSize(fir::unwrapSequenceType(cpnTy)))
return mlir::emitError(
loc, "fir.coordinate_of with a dynamic element size is unsupported");
- if (hasKnownShape || columnIsDeferred) {
+ if (shapeAnalysis->hasKnownShape || shapeAnalysis->columnIsDeferred) {
llvm::SmallVector<mlir::LLVM::GEPArg> offs;
- if (hasKnownShape && hasSubdimension) {
+ if (shapeAnalysis->hasKnownShape) {
offs.push_back(0);
}
+ // Else, only the column is `?` and we can simply place the column value
+ // in the 0-th GEP position.
+
std::optional<int> dims;
llvm::SmallVector<mlir::Value> arrIdx;
- for (std::size_t i = 1, sz = operands.size(); i < sz; ++i) {
- mlir::Value nxtOpnd = operands[i];
-
- if (!cpnTy)
- return mlir::emitError(loc, "invalid coordinate/check failed");
-
- // check if the i-th coordinate relates to an array
- if (dims) {
- arrIdx.push_back(nxtOpnd);
- int dimsLeft = *dims;
- if (dimsLeft > 1) {
- dims = dimsLeft - 1;
- continue;
- }
- cpnTy = mlir::cast<fir::SequenceType>(cpnTy).getElementType();
- // append array range in reverse (FIR arrays are column-major)
- offs.append(arrIdx.rbegin(), arrIdx.rend());
- arrIdx.clear();
- dims.reset();
+ int nextIndexValue = 1;
+ for (auto index : coor.getIndices()) {
+ if (auto intAttr = llvm::dyn_cast<mlir::IntegerAttr>(index)) {
+ // Addressing derived type component.
+ auto recordType = llvm::dyn_cast<fir::RecordType>(cpnTy);
+ if (!recordType)
+ return mlir::emitError(
+ loc,
+ "fir.coordinate base type is not consistent with operands");
+ int fieldId = intAttr.getInt();
+ cpnTy = recordType.getType(fieldId);
+ offs.push_back(fieldId);
continue;
}
- if (auto arrTy = mlir::dyn_cast<fir::SequenceType>(cpnTy)) {
- int d = arrTy.getDimension() - 1;
- if (d > 0) {
- dims = d;
- arrIdx.push_back(nxtOpnd);
- continue;
+ // Value index (addressing array, tuple, or complex part).
+ mlir::Value indexValue = operands[nextIndexValue++];
+ if (auto tupTy = mlir::dyn_cast<mlir::TupleType>(cpnTy)) {
+ cpnTy = tupTy.getType(getConstantIntValue(indexValue));
+ offs.push_back(indexValue);
+ } else {
+ if (!dims) {
+ if (auto arrayType = llvm::dyn_cast<fir::SequenceType>(cpnTy)) {
+ // Starting addressing array or array component.
+ dims = arrayType.getDimension();
+ cpnTy = arrayType.getElementType();
+ }
+ }
+ if (dims) {
+ arrIdx.push_back(indexValue);
+ if (--(*dims) == 0) {
+ // Append array range in reverse (FIR arrays are column-major).
+ offs.append(arrIdx.rbegin(), arrIdx.rend());
+ arrIdx.clear();
+ dims.reset();
+ ...
[truncated]
|
tblah
approved these changes
Feb 27, 2025
|
Thanks for the reviews! |
jeanPerier
added a commit
to jeanPerier/llvm-project
that referenced
this pull request
Feb 28, 2025
After llvm#127231, fir.coordinate_of should directly carry the field. I updated the lowering and codegen tests in 12731, but not the FIR to FIR tests, which is what this patch is cleaning up.
cheezeburglar
pushed a commit
to cheezeburglar/llvm-project
that referenced
this pull request
Feb 28, 2025
This patch updates fir.coordinate_op to carry the field index as attributes instead of relying on getting it from the fir.field_index operations defining its operands. The rational is that FIR currently has a few operations that require DAGs to be preserved in order to be able to do code generation. This is the case of fir.coordinate_op, which requires its fir.field operand producer to be visible. This makes IR transformation harder/brittle, so I want to update FIR to get rid if this. Codegen/printer/parser of fir.coordinate_of and many tests need to be updated after this change.
searlmc1
pushed a commit
to ROCm/llvm-project
that referenced
this pull request
Mar 2, 2025
lands and reverts: needs flang team to reland a8db1fb [flang] update fir.coordinate_of to carry the fields (llvm#127231
vzakhari
added a commit
to vzakhari/llvm-project
that referenced
this pull request
Mar 2, 2025
The issue came up after llvm#127231, when fir.coordinate_of, fed into a declare, only has the field attribute and no coordinates.
jeanPerier
added a commit
that referenced
this pull request
Mar 3, 2025
vzakhari
added a commit
that referenced
this pull request
Mar 3, 2025
The issue came up after #127231, when fir.coordinate_of, fed into a declare, only has the field attribute and no coordinates.
searlmc1
pushed a commit
to ROCm/llvm-project
that referenced
this pull request
Mar 3, 2025
Flang team looking into: land and reverted 3/1/25 a8db1fb [flang] update fir.coordinate_of to carry the fields (llvm#127231 subsequent landed and reverted, relates to prev. 9805854 [flang][NFC] clean-up fir.field_index legacy usages in tests (llvm#129219)
searlmc1
pushed a commit
to ROCm/llvm-project
that referenced
this pull request
Mar 3, 2025
This patch updates fir.coordinate_op to carry the field index as attributes instead of relying on getting it from the fir.field_index operations defining its operands. The rational is that FIR currently has a few operations that require DAGs to be preserved in order to be able to do code generation. This is the case of fir.coordinate_op, which requires its fir.field operand producer to be visible. This makes IR transformation harder/brittle, so I want to update FIR to get rid if this. Codegen/printer/parser of fir.coordinate_of and many tests need to be updated after this change.
searlmc1
pushed a commit
to ROCm/llvm-project
that referenced
this pull request
Mar 3, 2025
…9219) After llvm#127231, fir.coordinate_of should directly carry the field. I updated the lowering and codegen tests in llvm#12731, but not the FIR to FIR tests, which is what this patch is cleaning up.
searlmc1
pushed a commit
to ROCm/llvm-project
that referenced
this pull request
Mar 3, 2025
jph-13
pushed a commit
to jph-13/llvm-project
that referenced
this pull request
Mar 21, 2025
…9219) After llvm#127231, fir.coordinate_of should directly carry the field. I updated the lowering and codegen tests in llvm#12731, but not the FIR to FIR tests, which is what this patch is cleaning up.
jph-13
pushed a commit
to jph-13/llvm-project
that referenced
this pull request
Mar 21, 2025
The issue came up after llvm#127231, when fir.coordinate_of, fed into a declare, only has the field attribute and no coordinates.
stefankoncarevic
referenced
this pull request
in stefankoncarevic/rocMLIR
Mar 26, 2025
…482a
e60d3bf8482a [Hipcc] Mark ~HipBinBase as virtual (#1258)
bf3cb2786706 [Hipcc] Mark ~HipBinBase as virutal
158253a61e73 Remove symlink runtimes/cmake/Modules/FindLibcCommonUtils.cmake (#1255)
9889339a7059 Remove symlink runtimes/cmake/Modules/FindLibcCommonUtils.cmake
725843b5ce2b [device-libs] Allow link to llvm shared library for current distros (#1223)
595d6e244661 [IR] Find types used by DIExpressions in debug records (#892)
18e603698cc3 Revert "[OpenMP] Default to single SDMA engine" (#1232)
a4d4a48e80f7 merge main into amd-staging (#1231)
ad7727ec9490 merge main into amd-staging
907a11e32528 Revert "[OpenMP] Default to single SDMA engine"
71c8c003a634 Revert "Revert "[RemoveDIs] Enable direct-to-bitcode writing by default"" (#1229)
297f6d9f6b21 [libc++] Fix check for _LIBCPP_HAS_NO_WIDE_CHARACTERS in features.py (#131675)
ed19620b8c93 [VPlan] Make VPReductionRecipe a VPRecipeWithIRFlags. NFC (#130881)
00cad3ed2280 [SDAG] Handle extract_subvector in isKnownNeverNaN (#131581)
a5107be0317a [NFC][AMDGPU][GlobalISel] Make LLTs constexpr (#131673)
4cb1430c1cc6 [mlir][spirv] Fix a crash in `spirv::ISubOp::fold` (#131570)
745e16753fc1 [JSON][NFC] Move print method out of NDEBUG || DUMP (#131639)
d9c65af62654 [MLIR][GPUToNVVM] Support 32-bit isfinite (#131699)
d1156fcb5689 Revert "[libc++] Optimize num_put integral functions" (#131613)
50f8adb5c018 [RISCV] Accept '0(reg)' in addition to '(reg)' ifor vl1r.v/vl2r.v/vl4r.v/vl8r.v
173ce104bc98 merge main into amd-staging (#1230)
cb1d640b037b [clang][DepScan] resolve dangling reference to lambda that goes out of scope.
a2fbc9a8e3d3 [DirectX] Start the creation of a DXIL Instruction legalizer (#131221)
092e25571c09 AMDGPU: Add REQUIRES: asserts to machine pass violation test
94426df66a8d [compiler-rt][Darwin][x86] Fix instrprof-darwin-exports test (#131425)
e5ec7bb21b30 [flang][cuda] Set correct offsets for multiple variables in dynamic shared memory (#131674)
ad8f0e27606e [clang][DepScan] Pass references to ModuleDeps instead of ModuleID in lookupModuleOutput callbacks, NFCI (#131688)
c1fabd681fed [llvm][AMDGPU] Enable FWD_PROGRESS bit for GFX10+ (#128367)
541b8f2e14d4 [clang][driver] Use rva22u64_v as the default march for Fuchsia targets (#131183)
584f8cc30554 [clang][DependencyScanning] Track modules that resolve from "stable" locations (#130634)
5bf3f08cc967 [RISCV] Update some of the RVV memory ops in SiFive P400 & P600 sched models (#129575)
5f866666a6fb [CIR] Upstream initial support for unary op (#131369)
1b3164675777 [DFSan] Change placeholders from `undef` to `poison` (#131534)
ccf21094713e [Metadata] Change placeholder from `undef` to `poison` (#131469)
5b9006550d77 [libc++][NFC] Fix incorrect main() signatures
c3f750250a9b [MLIR][LLVM] Handle floats in Mem2Reg of memset intrinsics (#131621)
6d2b8285b3f5 [lldb] Support ordered patterns in lldbtest.expect (#131475)
166937b49dac [LV] Cleanup after expanding SCEV predicate to constant.
887cf1f8cea2 [AMDGPU][GlobalISel] Enable vector reductions (#131413)
128f381650aa [SPIR-V] Add `OpConstantCompositeContinuedINTEL` instruction (#129086)
74d4fc0a3ef0 [flang][cuda][NFC] Use ssa value for offset in shared memory op (#131661)
0191307bb258 [IR] Allow alignstack attribute on return values (#130439)
4336e5edbcc2 [SLP] Sort PHIs by ExtractElements when relevant (#131229)
6dbe82f061bf [NFC][DebugInfo] Wrap DILineInfo return type with std::optional to handle missing debug info. (#129792)
1c6208ac1576 merge main into amd-staging
2e6402ca2c6c [NFC] Add explicit initializer to PGOCtxProfReader's RootEntryCount
4ce1d1f1d90d [ADT] Add DenseSet::insert_range (#131567)
e3ef5f2928f4 [HLSL] Add bounds checks for the HLSL `fmod` vector arguments and return types (#131035)
6f659b0060d6 [clang][dataflow] For bugprone-unchecked-optional-access report range (#131055)
d2e1e3034801 [RISCV] Rename some DecoderNamespaces and cleanup debug messages. NFC (#131409)
9eb6b37e397e [RISCV] Put CV_ELW back in XCV DecoderNamespace.
bbaf743c461e [X86] Cleanup test coverage for #109272
20cdffbd2761 [X86] combineConcatVectorOps - extend VPERMILPD handling to support 512-bit types
24e88b0e6bc0 [libc++] Add remaining benchmarks from [alg.modifying.operations] (#127354)
911b200ce339 [Clang] Constant Expressions inside of GCC' asm strings (#131003)
d0a5845700a3 [IR] Find types used by DIExpressions in debug records
2443fe537f8b [gn build] Port af5abd9a682d
af5abd9a682d [HLSL] add extra scalar vector overloads for clamp (#129939)
279e82fca7f2 Revert f9146ccbe940d8b8eb15e7686a511a28eb0abc6b (#131656)
f6a7306beaf5 [clang][CIR] Add missing dependency on MLIR headers (#131057)
ead9d6a56d76 [SLP]Check VectorizableTree is not empty before accessing elements
cfa07ccdfcf0 [clang][bytecode] Fix builtin_memchr with non-0 start index (#131633)
ca1bde0b91a6 [clang][bytecode] Check dtor instance pointers for active-ness (#128732)
c53caae1d0ba [libc++][NFC] Remove dead link in comment
046041842022 [clang][driver][NFC] Remove else after return (#131182)
681b24132c5d [gn build] Port fbb8929c9d15
fbb8929c9d15 [lldb-dap] Updating RequestHandler to encode/decode arguments and response. (#130090)
47f7daab06e4 [libc][docs] Add glob implementation status doc and include in CMakeLists (#126923)
b3c5031b0739 [HLSL] Remove HLSLResource attribute (#130342)
9a92fe0f8d34 [X86] Add additional test coverage for #109272
561f1d0b7cbb [X86] add test coverage for concatenation to 512-bit VPERMILPD nodes
800593a01479 [MLIR][LLVM] Avoid duplicated module flags in the export (#131627)
67f1c033b8ff [VPlan] Remove createReduction. NFCI (#131336)
54cb4059da27 [X86][ISel][FMA] Get a handle on operand nodes when negating FMA (#130176)
6fcb486ba127 [device-libs] Allow link to llvm shared library for current distros
b00ad366323c [RISCV] Use hasFeature instead of checkFeature in llvm-exegesis. NFC (#131401)
eef5ea0c42fc [VPlan] Account for dead FOR splice simplification in cost model (#131486)
b24b984d092d merge main into amd-staging (#1219)
e2c43ba98162 [NFC][AMDGPU] Auto generate check lines for `llvm/test/CodeGen/AMDGPU/packed-fp32.ll` (#131629)
7054655c52e0 [llvm-exegesis] Add myself as an llvm-exegesis maintainer (#131580)
a4510aa7cb60 [flang-rt] replace the triple dir to 'aix' for flang-rt to be consistent with clang on AIX. (#130875)
3959bbc1345b [NFC] Remove trailing white spaces from `llvm/docs/LangRef.rst`
52e7ca9279b4 [LLVM][NVPTX] Add support for ldmatrix extensions introduced in PTX 8.6 (#124899)
269c40fafc80 [X86] Add tests for concatenation of VPERMV nodes
37c3fbfa5ae9 [X86] Add test showing failure to merge concatenatable VPERMV3 nodes
1b237198dc9d Reapply "[lldb] Implement basic support for reverse-continue (#125242)" (again) (#128156)
83356f3b62e9 [clang][bytecode] Compile functions lazily (#131596)
9455df969ef3 [Transforms] Avoid repeated hash lookups (NFC) (#131556)
e71686ed1539 [TargetParser] Avoid repeated hash lookups (NFC) (#131555)
8789c0083de1 [Transforms] Avoid repeated hash lookups (NFC) (#131554)
b8317df8d8f6 [FileCheck] Avoid repeated hash lookups (NFC) (#131553)
8bc0f879a052 [AMDGPU][True16][CodeGen] D16 LDS load/store pseudo instructions in true16 (#131427)
2ff370f45266 Warn about virtual methods in `final` classes (#131188)
1c3a9a853ce6 [AMDGPU] frame index elimination hit assertion for scavenged nonreg (#130287)
19adc69029ba [analyzer] Add [[maybe_unused]] forgotten in 57e36419b251 (#131617)
1e89a76a0490 [MLIR] Refactor to create vectorization convOp precondition check (#130181)
b79d53caaad7 [X86] X86MCInstLower.cpp - printConstant - don't assume the source constant data is smaller than the printed data
ed57ab0c2b8f [cmake] Move FindLibcCommonUtils to shared cmake, to fix standalone builds (#131586)
06546e005d2f [clang][NFC] Fix typo 'initializeation' (#131594)
175b862a3959 merge main into amd-staging
2a76734842c8 merge main into amd-staging (#1216)
17b4be8f63a9 [VPlan] Move setting name and adding VFs after recipe creation.(NFC)
e0223fa24aed [Clang][NFC] Rename SecondArgIsLastNamedArgument for clarity and consistency (#131346)
6085f3f6a80d [OpenMP] Address __kmp_dist_for_static_init issue (#129902)
f4feab927ba3 [NFC][KeyInstr] Add (LLVM_)EXPERIMENTAL_KEY_INSTRUCTIONS (cmake/)definition (#131344)
e57cd100ca29 Define LLVM_ABI and CLANG_ABI for __EMSCRIPTEN__ builds (#131578)
6eb32a2fa0d1 [clang] Change placeholder from `undef` to `poison` (#131533)
8c939f54b490 [WebAssembly] Change placeholder from `undef` to `poison` (#131536)
846cf86b2bbe libclc: Add missing gfx950 target (#131585)
0c34d7a9e7a4 [mlir][tosa] Require operand/result tensors of at least rank 1 for some operations (#131335)
5c73c5c9bf0e [X86][NFC] Add missing immediate qualifier to VSM3RNDS2 instruction (#131576)
93e0df07c2b3 [Flang][OpenMP] Allow zero trait score (#131473)
0878dd14b205 [AArch64][GlobalISel] Add coverage for arm64-neon-2velem.ll. NFC
e456579e346c [AMDGPU][RegBankCombiner] Add cast_of_cast and constant_fold_cast combines (#131307)
ab1dcac6db22 [AMDGPU][RegBankInfo] Promote scalar i16 and/or/xor to i32 (#131306)
27099982da2f [NFC][analyzer] Framework for multipart checkers (#130985)
7dcea28bf92e [AMDGPU] Add identity_combines to RegBankCombiner (#131305)
1f1f8200bdf5 AMDGPU: Switch simplifydemandedbits-recursion.ll to generated checks (#131317)
a6ae965cec96 AMDGPU: Switch scheduler-subrange-crash.ll to generated checks (#131316)
ee8a804cbab5 AMDGPU: Switch test to generated checks (#131315)
4f2ee07454b0 [BOLT][AArch64] Do not crash on authenticated branch instructions (#129898)
8cc6c2e80fb0 AMDGPU: Migrate more tests away from undef (#131314)
d9110858ee93 [Clang] Fix an incorrect assumption on getTemplatedDecl() (#131559)
a10e1e013597 [RISCV] Remove unused check prefixes from double maximum/minimum test. NFC
fd41f1b0ce1e [clang][analyzer] Add BugReporterVisitor messages for non-null fixed pointer (#129557)
6b47bba44087 [AMDGPU] Add intrinsics and MIs for ds_bvh_stack_* (#130007)
3af6c9fa832a [lldb][lldb-dap][NFC] Fix swapped logging directions for DAP messages. (#131544)
ccfabe838057 [clang-tidy-diff] Add an option to treat warnings as errors (#128221)
cdb355c60d36 merge main into amd-staging
e68153d050ce merge main into amd-staging (#1215)
9b1ce477af56 [gn build] Port 57e36419b251
57e36419b251 [analyzer] Introduce per-entry-point statistics (#131175)
c3f6d2c02496 [gn build] Port 3b1e18c2dba8
3b1e18c2dba8 [clang-tidy] Add new check bugprone-capture-this-by-field (#130297)
7eb8b7317847 [Flang][OpenMP][taskloop] Adding missing semantic checks in Taskloop (#128431)
78408fddccf3 [ExecutionEngine] Avoid repeated map lookups (NFC) (#131552)
05607a3f39f9 [CodeGen] Avoid repeated hash lookups (NFC) (#131551)
2c35cb6f16b2 [MC] Remove unneeded getNumFixupKinds
009d36222cfd [clang][CodeComplete] Add code completion for if constexpr and consteval (#124315)
6c867e27a7b5 [mlir] Use `getSingleElement`/`hasSingleElement` in various places (#131460)
f402953339fa [ORC] Fix code example in comment: SPS function sigs are function types. NFCI.
f75d75b8899a AMDGPU: Use MFPropsModifier modifier in SIFoldOperands (#127752)
8a1b4d0ed2bb [MC] Rework AVR #121498 to not add extra argument to shouldForceRelocation
b09b9ac1081d [llvm][CodeGen] Fix the empty interval issue in Window Scheduler (#129204)
e0fee65b8787 [PowerPC] Use Register in FastISel. NFC
c5a491e9ea22 [SCEV] Check whether the start is non-zero in `ScalarEvolution::howFarToZero` (#131522)
baab447aadd5 [llc] Report error in lieu of warning for invalid cl option (#128846)
de60c0e034f9 [MC] .reloc: move FirstLiteralRelocationKind check to evaluateFixup
687c9d359ee1 [CodeGen][NPM] Port FEntryInserter to NPM (#129857)
2a2d6d61b178 [RISCV] Remove unused check prefixes from half arith strict test. NFC
752aa81c4ff7 [clang][RISCV] Rename variable name in SemaRISCV. NFC (#131261)
e24e523150c2 [LoopVectorize] Add test for follow-up metadata for loops (NFC) (#131337)
95adcf28cea6 merge main into amd-staging
4fde8c341f91 [flang][cuda] Lower CUDA shared variable with cuf.shared_memory op (#131399)
02df29a45c3e Work around bitcode incompatibility with upstream (#1199)
4b86a7f3860a [clang-format] Update the minimum python version requirement
2e78abe788c5 [MSP430] Delete unneeded fixupNeedsRelaxationAdvanced
125c4db73018 [lldb][lldb-dap] setVariable request should send the correct response (#130773)
2dc123b33d51 [clang][opencl] Allow passing all zeros to reqd_work_group_size (#131543)
81ba00629668 [X86] nocf_check: disable tail call
91328dbae986 [clang-format] Correctly annotate user-defined conversion functions (#131434)
950bc6cd7745 [LoopFuse] Change placeholder from `undef` to `poison` (#131535)
82975d4ab962 merge main into amd-staging (#1211)
3e6f618e86f5 [llvm][ADT] Add `getSingleElement` helper (#131508)
40b703421377 [LV] Add tests for vector backedge elimination with early-exit loops.
ee29e16135a9 [LV] Reorganize tests for narrowing interleave group transform.
5ddd27c5e16d merge main into amd-staging
b6485765284a [CodeGen] Avoid repeated hash lookups (NFC) (#131495)
4e841d7d63cd [libc] add uefi docs (#131426)
7c98cddc5add [mlir] Expose `AffineExpr.shift_dims/shift_symbols` through C and Python bindings (#131521)
93ce34550451 [PowerPC] Avoid repeated hash lookups (NFC) (#131498)
1bc2108c49f4 [Transforms] Avoid repeated hash lookups (NFC) (#131497)
48ecec20a204 [Hexagon] Avoid repeated hash lookups (NFC) (#131496)
e2438ce94002 [BPF] Avoid repeated map lookups (NFC) (#131494)
8705e489dc17 [lldb] Remove use of comma operator (NFC) (#131233)
1ceaecdafebe merge main into amd-staging (#1209)
9829d457ae19 [libc++][test] Adds a test for a reserved name. (#131363)
616d1046c3e2 [compiler-rt][rtsan] fix ioctl interception for musl. (#131464)
215c0d2b651d [AArch64][GlobalISel] Reorder getActionDefinitionsBuilders. NFC
d928a671b84a [mlir][Vector] Refactor VectorEmulateNarrowType.cpp (#123529)
4e9894498e16 [VPlan] Truncate VFxUF if needed in VPWidenPointerInduction::execute.
dc9a183ac6aa [NFC][Cloning] Move DebugInfoFinder decl closer to its place of usage (#129154)
3a106c9092f7 merge main into amd-staging
926d980017d8 [AArch64][GlobalISel] Some minor reordering of types for consistency. NFC
a9ab5fd079f2 merge main into amd-staging (#1205)
90e6ba606f7a [AArch64][GlobalISel] Remove min/max v2s64 clamp
3fac23505ffd [AArch64][GlobalISel] Add cttz, ctlz and ctpop test coverage. NFC
de03e102d1ea [DirectX] Change placeholders from `undef` to `poison` used in fully instantiated vector [NFC] (#130970)
f4043f451d0e Skip more WebKit checker tests on targets where builtin is not supported. (#131501)
1e02442df6cf [Github][Docs] Add best practice for top level read permissions (#131470)
3fe914c9faa8 [X86] Use Register and MCRegister. NFC
0689d23ab308 [C++20][Modules] Prevent premature calls to PassInterestingDeclsToConsumer() within FinishedDeserializing(). (#129982)
508db53d1af5 [AMDGPU] Avoid repeated hash lookups (NFC) (#131493)
d0177670a0e5 [ADT] Avoid repeated hash lookups (NFC) (#131418)
f5f8f3c64929 [GSYM] Update gSym unit test with stable / portable path (#131204)
f7cab64684bc [libc++][numeric][NFC] Cleanup *Saturation arithmetic* tests (#101826)
e30a5d657034 [libc++][NFC] Simplify string a bit (#127135)
2091547d4cc9 [PPC codegen test] NFC: Fix RUN line; fix DATA checks to match 64-bit
4286f4dccee9 [AArch64][GCS][LLD] Introduce -zgcs-report-dynamic Command Line Option (#127787)
247f995e1b42 merge main into amd-staging
7722d7519ca2 [MC] evaluateAsRelocatableImpl: remove the Fixup argument
ecadccf07789 merge main into amd-staging (#1204)
ff2ed154a8a9 [MC] evaluateAsAbsolute requires MCValue::RefKind==0
5e65b40f9c0f [clang-tidy] detect explicit casting within modernize-use-default-member-init (#129408)
d6fbffa23c84 [MC] evaluateAsRelocatable: remove the Fixup argument
0bd8a75a0c8c [VPlan] Fix formatting after 6a8d5f22f.
6a8d5f22ffc5 [VPlan] Don't access canonical IV in VPWidenPointerInduction::execute.
79d84a878e83 MipsMCExpr: remove unneeded folding and fix a crash for %hi(und-$L3)
9a1e39062b2a [Clang] Do not emit nodiscard warnings for the base expr of static member access (#131450)
aadfa9f6c872 [LV] Add additional tests for narrowing interleave groups.
6616acd80cd9 [Github] Change to step-security fork of changed actions
7c26407a20b5 [LLD][COFF] Clarify EC vs. native symbols in diagnostics on ARM64X (#130857)
8560da28c69d [PowerPC] Simplify PPCMCExpr::evaluateAsRelocatableImpl
fc6fd6a2f119 [NFC][Cloning] Clean up comments in CloneFunctionInto (#129153)
f7b3eadfd524 merge main into amd-staging
316811060775 [AddressSanitizer] Remove memory effects from functions (#130495)
37a57ca257c1 [FMF] Set all bits if needed when setting individual flags. (#131321)
5f449b9a5dae [RISCV] Allow RISCVMCExpr folding in the absence of linker relaxaxation
56b05a0d6ba6 [VPlan] Use VFxUF in VPWidenPointerInductionRecipe.
5ae8f25dcc89 [RISCV] Move fixELFSymbolsInTLSFixups to getRelocType
a1a29c3cb32a [Mips] Move fixELFSymbolsInTLSFixups to getRelocType
8ff27bbd8449 [VPlan] Remove unneeded select in VPWidenPointerInductionRecipe (NFC).
6c2f8476e7e7 [mlir][Transforms] Dialect Conversion: Add 1:N support to `remapInput` (#131454)
911953a2e408 [Sparc] Move fixELFSymbolsInTLSFixups to getRelocType
f4ea1055ad57 [SystemZ] Implement i128 funnel shifts
4155cc0fb319 [SystemZ] Recognize carry/borrow computation
4a4987be360a [SystemZ] Optimize vector zero/sign extensions
cdc786498650 [SystemZ] Optimize widening and high-word vector multiplication
7af3d3929e85 [SystemZ] Optimize vector comparison reductions
86ae25d2be59 [CodeGen][X86] Use Register in TTI unfoldMemoryOperand interface. NFC
fc17114ac00d [NFC][Cloning] Remove now unused CollectDebugInfoForCloning (#129152)
5bc8db884e3a [libc++][NFC] fix exception_guard include guard name (#131370)
d1700cdbf274 [ExecutionEngine] Avoid repeated hash lookups (NFC) (#131423)
6a1fd24e9ac0 [CodeGen] Avoid repeated hash lookups (NFC) (#131422)
03205121d264 [Analysis] Avoid repeated hash lookups (NFC) (#131421)
4230858a52ab [ARM] Avoid repeated map lookups (NFC) (#131420)
f83726e6add0 [profile] Use fprofile-continuous in compiler-rt tests (#126617)
65e68a30787d [lldb] Update dwim-print to show expanded objc instances (#117500)
e2c96a0ae30e merge main into amd-staging (#1203)
d781ac1cf0d5 [C23] Add __builtin_c23_va_start (#131166)
5cc2ae0b5190 [libc++][test] Skip a `is_virtual_base_of` test for apple-clang-17 (#131438)
1e6ba3cd2fe9 [NFC][Cloning] Remove now unused FindDebugInfoToIdentityMap (#129151)
60434bf2e2a3 merge main into amd-staging
6a030b300547 [VPlan] Remove unused VPlanIngredient (NFC).
f444c816c7ed merge main into amd-staging (#1202)
94dc397c7e1a [bazel] Add missing dependencies for fd24805c8e67c921991e82463bdc23563caf744e
254951749feb [X86][APX] Remove the EFLAGS def operand rather than the last one (#131430)
3b5413c77fd5 [CodeGen] Use MCRegister in DbgVariableLocation. NFC
6c66cda40d7c [AntiDepBreaker] Use MCRegister. NFC
720730041dee [MC] Remove empty fixELFSymbolsInTLSFixups overrides
2ada0c1e6163 [AArch64] Move fixELFSymbolsInTLSFixups to getRelocType
aead088f02b9 [AMDGPU] Avoid repeated hash lookups (NFC) (#131419)
1762f16f6cc4 [InstCombine] Fold `umax/umin(nuw_shl(z, x), nuw_shl(z, y)) -> nuw_shl(z, umax/umin(x, y))` and `umax/umin(nuw_shl(x, z), nuw_shl(y, z)) -> nuw_shl(umax/umin(x, y), z)` (#131076)
215c47e4d3b2 [ctxprof] Missing test update post #131201 (#131428)
e642ed8b4ac7 [AVR] Remove initializeVariantKinds
a4976ca67498 Move Hexagon-specific MCSymbolRefExpr::VariantKind to HexagonMCExpr
b034905c8218 [ctxprof] Capture sampling info for context roots (#131201)
a3ce1cc3b0ea [MC] Remove unused VK_AVR_*
221bc55806db Move AVR-specific MCSymbolRefExpr::VariantKind to AVRMCExpr
bb694998d2f2 [libc] init uefi (#131246)
e86081b6c275 [flang][cuda] Convert cuf.shared_memory operation to LLVM ops (#131396)
4fb20b85fd81 [flang][cuda] Compute offset on cuf.shared_memory ops (#131395)
2147518d832d merge main into amd-staging
4b1b629d6034 [BOLT] Fix a warning
bac21719a8a9 [BOLT] Pass unfiltered relocations to disassembler. NFCI (#131202)
d52ec1e9ddc6 [MLIR][NFC] fix msvc debug build errors (#131393)
5265412c13b4 [MLIR][LLVMIR] Import: add flag to prefer using unregistered intrinsics (#130685)
29a000023caf [MLIR][LLVMIR] Add module flags support (#130679)
3be3c6130c66 Work around bitcode incompatibility with upstream
3de1b9e56c35 [RegAllocFast] Add missing source lines. (#1188)
b7852939b592 [NFC][AMDGPU] Replace multiple calls to `MI.getOpcode()` with `Opcode` (#131400)
f326036767aa [flang-rt] Added IsContiguousUpTo runtime function. (#131048)
b4f5dcc65a36 [lldb-dap] Reword the description in package.json
c8b8415b1af1 [flang-rt] Install flang_rt.cuda with the toolchain. (#131373)
5b2a8819fbe0 Revert "[Transforms] LoopIdiomRecognize recognize strlen and wcslen (#108985)" (#131405)
bf6357f0f51e [Transforms] LoopIdiomRecognize recognize strlen and wcslen (#108985)
89889149cd21 [lld-macho] Improve ICF thunk folding logic (#131186)
4818623924a6 [flang][cuda] Add cuf.shared_memory operation (#131392)
471f034b1cda [clang-tidy][NFC] clean ReleaseNotes.rst (#130626)
a862b6deae98 [flang][cuda] Lower shared global to the correct NVVM address space (#131368)
c5ef332bbbe7 Manual update of LLVM_MAIN_REVISION to 530618 (#1198)
e525567755d0 [libc] Make RPC server handling header only (#131205) (#1192)
f32942263411 merge main into amd-staging (#1196)
fbf0276b6a7a [SLP] Reorder reuses mask, if it is not empty, for subvector operands
dccc0a836c20 [NFC][AMDGPU] Replace more direct arch comparison with isAMDGCN() (#131379)
605a9f590d91 [SLP]Check if user node is same as other node and check operand order
8413f4d837a9 [llvm-objcopy] Apply encryptable offset to first segment, not section (#130517)
8e4c4999cce6 Manual update of LLVM_MAIN_REVISION to 530618
456963de9690 [mlir] Fix warnings
6ac63129b7ef [LLVM] Only enable `-fno-lifetime-dse` in LTO mode (#131381)
78f74f686bfe [Darwin][ASan][Test] Create a noinlined wrapper function for reliable suppression in test. (#131247)
cb64a363ca71 [HLSL] Make sure `isSigned` flag is set on target type for `TypedBuffer` resources with signed int vectors (#130223)
f33384184462 [RISCV] Add nf argument to VReg class instead overriding with a let. NFC (#131235)
9c86198caf36 [SLP] Update vector value for incoming phi node, beeing vectorized already
955c02dc9c9c [mlir][tosa] Check for compile time constants in the validation pass (#131123)
7a7c33d4f021 [lldb] Sort source files alphabetically in API/CMakeLists.txt (NFC)
2490f7f07673 [mlir][Linalg] Allow expand shape propagation across linalg ops with dynamic shapes. (#127943)
fd24805c8e67 Reapply [mlir][xegpu] Add XeGPU subgroup map propagation analysis for XeGPU SIMT distribution. (#131380)
8bceb777a140 [NFC][Coro] Remove now unused CommonDebugInfo in CoroSplit (#129150)
b17af9d8ee4e [NFC][llvm/IR] comparison of unsigned expression in ‘>= 0’ is always true (#130843)
0f1175c8ad0d [AMDGPU][True16][CodeGen] enable true16 for more codegen test patch 3 (#131212)
b1fe7dabceec [AMDGPU][True16][CodeGen] enable true16 for more codegen test patch 2 (#131210)
0b688f3ce154 [AMDGPU][True16][CodeGen] enable true16 for more codegen test patch 1 (#131206)
f798c7625e56 merge main into amd-staging
7598ceaf51e0 [gn build] Port c84d8e8f1c40
c84d8e8f1c40 [clang][modules] Introduce new `ModuleCache` interface (#131193)
d0a0de50f7dc [libc] Fix implicit conversion warnings in tests. (#131362)
51c706c11955 [NFC][AMDGPU] Replace direct arch comparison with `isAMDGCN()` (#131357)
2f9d94981c0e [BOLT] Change Relocation Type to 32-bit NFCI (#130792)
e9fc7683a54d [CIR] Upstream basic support for sizeof and alignof (#130847)
3ec693d3529c [X86] combineConcatVectorOps - use ConcatSubOperand for X86ISD::VPERMV operand concatenation (#131352)
43c34e38d496 [Comgr][NFC] Fix typo in test error log (#1181)
f8ae24a916e7 [Comgr][NFC] Add space (#1133)
e6382f211135 SelectionDAG: neg (and x, 1) --> SIGN_EXTEND_INREG x, i1 (#131239)
3fcd921aa492 Revert "[mlir][xegpu] Add XeGPU subgroup map propagation analysis for XeGPU SIMT distribution." (#131364)
352b9e65be17 [X86] combineConcatVectorOps - extend ISD::ROTLI/VROTRI handling to support 256-bit types
3f62718c4a90 AMDGPU: Migrate some tests away from undef (#131277)
5eb557774df6 [mlir][xegpu] Add XeGPU subgroup map propagation analysis for XeGPU SIMT distribution. (#130240)
87a55191a5c5 MIR: Replace undef with poison in some MIR tests (#131282)
217fc6579bd0 [lldb][lldb-dap] Return optional from json utils (#129919)
86329ba4d904 [HLSL] Remove old resource annotations (#130338)
a8949b16e64f [X86] Add test coverage showing failure to concatenate matching rotate nodes
2ee7ba47dc97 [RISCV] Move Size and CopyCost overrides for vector register to the VReg class. NFC (#131222)
cbbcc3d13b77 [C2y] Claim conformance to WG14 N3460 (#131196)
beb4a482973d [InstCombine] Use known bits to simplify mask in foldSelectICmpAnd (#128741)
cd54d581b506 [AMDGPU][True16][CodeGen] add v_cndmask_t16 to hazardmask (#128912)
3ac5d8da615e [mlir-lsp] Abstract input and output of the `JSONTransport` (#129320)
b936ef18559a [RISCV] Reorder include of RISCVInstrInfoZi* before C and Zc*. NFC (#131274)
8c31bb7da34a [GSYM] Fix incorrect comparison in gSYM creation (#131197)
bdb4012fe3f9 [CodeGen] Remove parameter from LiveRangeEdit::canRematerializeAt [NFC]
befb52db94cc [Clang] Remove use of 'temporary' toolchains for offload deduction (#131332)
999700ca477d [libc++][test] Skip some `is_implicit_lifetime` tests for apple-clang-17 (#131302)
da3ee9763266 StandardInstrumentation: Fix -ir-dump-directory with -print-before-pass-number (#130983)
792a6f811988 [RemoveDIs] Remove "try-debuginfo-iterators..." test flags (#130298)
42ddb5501748 [X86] combineINSERT_SUBVECTOR - peek through bitcasts to find a concatenation of subvector shuffles (#131331)
eeb27331dce4 [flang-rt] Use --as-needed for linking flang-rt libraries. (#130856)
fdb4b89bc0e5 [libc] Fix memmove macros for unreocognized targets
00f9c855fb3b [flang] Added fir.is_contiguous_box and fir.box_total_elements ops. (#131047)
52cd27e60b24 Revert "[Aarch64] [ISel] Don't save vaargs registers if vaargs are unused (#126780)"
72b8744aa530 [MLIR][OpenMP] Reduce overhead of target compilation (#130945)
f2541cee44ed [clang] Remove usage of llvm-spirv in clang LIT tests (#131158)
fbc3f7b6bdd3 [OpenMP][Offload][AMDGPU] Add envar for setting CU multiplier (#1143)
e9384896aec5 merge main into amd-staging (#1189)
e8117026a979 [libc] Default to `byte_per_byte` instead of erroring (#131340)
244cf89f143a [mlir][python] Small optimization to mlirApiObjectToCapsule. (#131160)
841f4637a078 [NFC][Coro] Use CloneFunctionInto for coroutine cloning instead of CloneFunction<Part> (#129149)
389ed474e81f [BPF] Remove unnecessary BitCast operations (#131260)
2198f546b167 [BPF] Avoid repeated hash lookups (NFC) (#131265)
2a09523480fc [clang] Add diagnostic for unresolved using declaration that shadows template parameters (#131328)
19b25a452439 [dsymutil] Avoid repeated hash lookups (NFC) (#131268)
1b42be6fe8b8 [Utils] Avoid repeated hash lookups (NFC) (#131267)
8d333e167503 [SelectionDAG] Avoid repeated hash lookups (NFC) (#131266)
4acfeafd8fa4 [Tooling] Avoid repeated hash lookups (NFC) (#131264)
0f98d1b9fa56 [CIR] Don't generate ClangIR after an unrecoverable error occured (#130971)
8f011e2c7f00 [Sema] Avoid repeated hash lookups (NFC) (#131263)
bbd1bb40573d [SLP]Set insert point for split node with non-scheulable instructions after the last instruction
9387281b6e98 [msan][NFCI] Add test for llvm.x86.vcvtph2ps.{128,256} (#131244)
b43620661d16 [clang] Fix inaccurate wording of warn_second_arg_of_va_start_not_last_named_param (#131238)
05df923b0e8a [CI] Add dateutil dependency to the metrics container (#131333)
737a0aeb6b4e [NFC][PowerPC] cleaned dead code of PPC.cpp and PPC.h (#130994)
a0b175cb348b [SimplifyCFG] Treat `extract oneuse(op.with.overflow),1` pattern as a single instruction (#128021)
31a4f388dac7 [libc] Make RPC server handling header only (#131205)
1a68269e2805 [clang-tidy] support pointee mutation check in misc-const-correctness (#130494)
20b7f5982622 [Clang] [Tests] Add some more tests around non-local/non-variable declarations in C for loops (#131199)
c979ce7e3629 Add IRBuilder::CreateFMA (#131112)
e4a8969e5657 [clang-tidy] Avoid processing declarations in system headers (#128150)
6962cf1700c2 Rename ExpandLargeFpConvertPass to ExpandFpPass (#131128)
73e93ec3a2ab [X86] combineConcatVectorOps - attempt to recursively call combineConcatVectorOps from inside ConcatSubOperand. (#131303)
80079c9c2f55 [libc] Fix new warning in DyadicFloat::as_mantissa_type_rounded (#131148)
e45090e5f0bf Fix BUILD.bazel due to a16c225
a54f75b46b10 [AArch64][GISel] Silence warning caused by #130156
ec95ce358c88 [lldb][AIX] Added base files for NativeProcess Support for AIX (#118160)
11d35a0a94cf AMDGPU: Use generated checks in coalescer_distribute.ll (#131276)
a9843ac28540 AMDGPU: Use generated checks in unchecked test (#131275)
3b6d0093aa82 [LV][NFC] Refactor code for extracting first active element (#131118)
a58a6a95b078 [SPIR-V] Support SPV_INTEL_fp_max_error extension for `!fpmath` metadata (#130619)
9af538420aca [AArch64][GISel] Fix lowering of fp16 intrinsics (#130156)
fb0e77ae6293 merge main into amd-staging
44f4e43b4fbb [CI] Extend metrics container to log BuildKite metrics (#130996)
6e17a6c1951b Add missing source lines.
0a5847f1c144 [X86] combineConcatVectorOps - pull out repeated getOpcode() calls. NFC.
8b217ebdd2bd [OPENMP] [FLANG] add support for formatted Io which generates call to… (#1185)
3ff3b29dd624 [flang] lower remaining cases of pointer assignments inside forall (#130772)
7bae61370d61 AMDGPU: Use generated tests in reg-coalescer-sched-crash.ll test (#131259)
4c1a1a5bf3b9 AMDGPU: Add generated checks to compile only test (#131258)
75cf046f0306 Re-apply "[AMDGPU] Fix test failures when expensive checks are enabled (#130644)"
fc8b2bf2f859 [MLIR][LLVM] Import dereferenceable metadata from LLVM IR (#130974)
bddf24ddbdb2 [Flang] Add omp_lib dependency to check-flang (#130975)
9b83ffb5c609 AMDGPU: Switch a test to generated checks which only tested labels (#131257)
8d0205f6ddae AMDGPU: Remove undef in subreg-coalescer-crash.ll (#131256)
3529c64dbc99 AMDGPU: Switch a test with only function label checks to generated (#131255)
ac94ccd56499 AMDGPU: Replace undef references with poison in some MIR tests (#131254)
17eb7777c0ff AMDGPU: Use generated checks in test missing checks (#131110)
26324bc1bf39 [VPlan] Move FOR splice cost into VPInstruction::FirstOrderRecurrenceSplice (#129645)
b6370f5a04a6 merge main into amd-staging (#1186)
f2e10278efe7 [lldb] Support discontinuous functions in another Disasembler overload (#130987)
77ad06192341 [OpenMP] Update OpenMP runtime to adopt taskgraph clause from 6.0 Specs (#130751)
467ad6a03583 [clang-format] Add support for absl nullability macros (#130346)
6b7daf224933 [MachineCombiner][Targets] Use Register in TII genAlternativeCodeSequence interface. NFC (#131272)
2a48995a03ca [ARM] Pass ArrayRef by value instead of const reference. NFC
8727097ffd01 [RISCV][Sema] Add feature check for target attribute to VSETVL intrinsics (#126064)
87916f8c32eb [CodeGen][NPM] Port MachineBlockPlacement to NPM (#129828)
42748a454fc9 [Clang driver] Diagnose `-maix-shared-lib-tls-model-opt` on wrong targets (#130865)
f34385dd1bac [AMDGPU][NPM] Port GCNCreateVOPD to NPM (#130059)
a5b95487d634 [ctxprof] Missing test for #131269 (#131271)
4a3ee4f72dcd AMDGPU: Make fma_legacy intrinsic propagate poison (#131063)
37706894f8bd AMDGPU: Make fmul_legacy intrinsic propagate poison (#131062)
a716459f2d0e AMDGPU: Make ballot intrinsic propagate poison (#131061)
75349d7ca4c6 Skip unretained-call-args.mm on platforms where builtin is not supported. (#131252)
0d8a22d6ad5d AMDGPU: Make fmed3 intrinsic propagate poison (#131060)
9b887f5277ec AMDGPU: Make cvt_pknorm and cvt_pk intrinsics propagate poison (#131059)
61adca7c9708 [ctxprof] Fix initializer in PGOCtxProfLowering (#131269)
03614b9a8a2f [BOLT] Workaround failures (#131245)
a6463f41135d [RISCV] Shrink the size of the VLMul field in RegisterClass target flags. Use uint8_t for TSFlags. NFC (#131227)
bde2636fb5fb [libc][NFC] Add the missing angle bracket in `wchar.h` (#131161)
7c63334c3f95 Revert "[libc] Make RPC server handling header only (#131205)"
910514c6ab32 AMDGPU: Replace some test undef uses with poison (#131103)
06c379a3498c AMDGPU: Replace more undef test pointer uses with poison (#131102)
37c8792e53c7 AMDGPU: Replace test uses of ptr addrspace(5) undef with poison (#131101)
ad993687b1a6 AMDGPU: Replace ptr addrspace(4) undef uses with poison in tests (#131095)
313abfbf1b57 merge main into amd-staging
482b93cabb01 Fix style
315c02aa023e [VPlan] Fix crash with inloop fmuladd reductions with blend (#131154)
0abbf1888f4e Address comments
f23bbf69f6ac [X86] Avoid repeated hash lookups (NFC) (#131069)
f4ee62c7c3ac [IPO] Avoid repeated hash lookups (NFC) (#131068)
e838ca1f49c2 [CodeGen] Avoid repeated hash lookups (NFC) (#131067)
ca641b29190f [Analysis] Avoid repeated hash lookups (NFC) (#131066)
7714df909054 [Sema] Avoid repeated hash lookups (NFC) (#131065)
dec5589d90f0 [AST] Avoid repeated hash lookups (NFC) (#131064)
9e538b312fbc merge main into amd-staging (#1180)
8437b7f55847 [libc] Make RPC server handling header only (#131205)
214d47af9bf9 [SandboxVec][PassManager][NFC] Fix PM printing
299cb5d88fad [LLVM] Update llvm-mca / MCA maintainers (#131187)
dcec224240d6 Lex: add support for `i128` and `ui128` suffixes (#130993)
3b5842c9c41a [ORC] Make runAllocActions and runDeallocActions asynchorous.
864a53b4a414 Reapply "Use global TimerGroups for both new pass manager and old pass manager timers" (#131173) (#131217)
d0c869521ce0 [Support] Prevent leaking unique lock files (#130984)
e26bcf1627cc [CodeGen] Use early return to simplify SplitEditor::defFromParent [NFC]
befa037c135c [mlir][affine] Guard invalid dim attribute in the test-reify-bound pass (#129013)
fc28f83bf5fa [SLP] Precommit test (#131236)
ab58a3c35b6f XFAIL malloc_zone.cpp for darwin/lsan (#131234)
55e5e749ce0c [RISCV] Remove some unnecessary hasSideEffects = 0 and sink some to their base class. NFC (#131044)
d1deaed0d284 [lldb] Split some lldb-server tests to avoid timeout (#129614)
8b9031f245a2 [lldb-dap] Support vscode launch URLs (#125843)
2a244bb839e3 Simplify Target::RunStopHooks() (#129578)
662bd4ca4714 [stdio][baremetal] Fix templating for print functions (#131232)
e0e80dbe4320 [Clang codegen][PPC] Produce AIX-specific "target features" only for AIX (#130864)
998511c8ef57 [debugserver] Fix mutex scope in RNBRemote::CommDataReceived (#131077)
08d15e3f64e3 Revert "reapply [llvm] add support for mustache templating language" (#131228)
54fa9e2aa46e [OPENMP] [FLANG] add support for formatted Io which generates call to API _FortranAioBeginExternalFormattedOutput
0ab9f78957b2 [docs] Fix link appearing improperly (#131225)
0f8c075f7c2c [llvm] Match llvm.type.checked.load.relative semantics to llvm.load.r… (#129583)
dfb661cd1c5f [LAA] Add extra tests for #128061.
c59c3874179f [RISCV] Improve Disassembler test xqci-invalid.txt (#130039)
96c02268ceaf [Comgr] Re-populate DataAction options for second compilation (#1130)
57d87ed7f097 [flang][NFC] Add parenthesis to avoid warning (#131219)
37d99e9c6e50 [GOFF] Refactor GOFFOstream (#131143)
4cc4a978faca [Comgr][NFC] Fix typo in test error log
e823449f66ac [lldb][debugserver] Synchronize interrupt and resume signals (#131073)
bbb244c0fdd3 [RISCV] Set CopyCost on register classes [NFC] (#131185)
688d7a5d0ade [X86] combineConcatVectorOps - extend ISD::VECTOR_SHUFFLE handling to support 512-bit types (#131189)
cfc3b220cb7f merge main into amd-staging
4bcf1e61eb84 Update the libc BUILD.bazel file with selects for Windows builds. (#131172)
1c4551474847 [mlir][tosa] Fix bug causing quantized pad const creation crash (#131125)
c30ff922caba [msan] Handle llvm.x86.vcvtps2ph.128/256 explicitly (#130705)
8c7f0eaa6ee3 Reduce memory usage in AST parent map generation by lazily checking if nodes have been seen (#129934)
8aa835c2b5a9 [ctxprof] Fix warnings post PR #130655 (#131198)
d642eec78fc9 [HEXAGON] Fix semantics of ordered FP compares (#131089)
64f67f870d6e [mlir][AMDGPU] Enable emulating vector buffer_atomic_fadd for bf16 on gfx942 (#129029)
b36bf47b2508 [LLDB][Telemetry]Init field to nullptr. (#131191)
ca4399036f59 [mlir][linalg] Add FoldReshapeWithGenericOpByCollapsing pattern (#131029)
f13d58303f8b [VPlan] Pass some functions directly to all_of (NFC).
933ecf5f30f1 [mlir] adds `[[maybe_unused]]` to variables that might not be used (#131184)
a16c225b4079 [mlir][xegpu] Convert Vector contraction to XeGPU (#122115)
6fea340023af [libc] Fix non-templated uses of `printf_core::Writer` (#131149)
02575f887bc7 [VPlan] Use VPInstruction for VPScalarPHIRecipe. (NFCI) (#129767)
e8e267ec42ed [libc] Remove use of C++ STL features from `rpc_server.cpp` (#131169)
3bd71cbec752 [libc++] Fix ambiguous call in {ranges, std}::find (#122641)
e61859f14ddd [RISCV] Add Qualcomm uC Xqcili (load large immediates) extension (#130012)
c2b66ce655e5 [flang][OpenMP] Silence unused-but-set-variable message (NFC) (#130979)
4baf1c03fa94 [libc++] Optimize ranges::rotate for vector<bool>::iterator (#121168)
369da8421c2f [flang][cuda] Allow assumed-size declaration for SHARED variable (#130833)
e93e0dd10cda [NFC] Fix formatting for #80963 (#131100)
5af5fb055527 [C2y] Claim conformance to WG14 N3482
d8dfdafc1d75 [Support] Introduce new `AdvisoryLock` interface (#130989)
bd0d28ac257d [CIR] Upstream basic support for ArrayType (#130502)
31ebe6647b7f Revert "Use global TimerGroups for both new pass manager and old pass manager timers" (#131173)
aa612f3ade66 [NFC][Cloning] Replace DIFinder usage in CloneFunctionInto with a MetadataPredicate (#129148)
09d8e442ac28 [llvm][Timer] Use global TimerGroups for both new pass manager and old pass manager timers (#130375)
09a36c82793b [lldb][NFC] Correct whitespace in SearchForKernelWithDebugHints
65220fc5aec9 [PPC] Fix coding style violations in PPCTargetTransformInfo.cpp (#130666)
7addc3557e2d [Comgr][Cache] Fix broken test: spirv-translator-cached.cl (#1151)
d06937aea38a [Flang][NFC] Fix typo (#130960)
cb0d9fcd72be [HLSL][NFC] Update HLSL AST tests to be more readable (#130910)
56d61ed04938 merge main into amd-staging (#1167)
b3b007078212 Follow up to #130230 - update missed test
d2d052b3e665 [DirectX] Use "texture" not "SRV" when pretty printing resources (#130230)
96637b46f156 [Clang] Improve `getReplacedTemplateParameterList()` const correctness (#131165)
15a5b3a19216 [AMDGPU][True16][CodeGen] gisel true16 for ICMP (#128913)
c9d7f707c101 [Headers][NFC] Deduplicate gpu_match_ between targets via inlining (#131141)
2044dd07da99 [InstrProf] Remove -forder-file-instrumentation (#130192)
0ed5f9b22bcb [MLIR] NFC. Fix unused warning in affine loop utils
01aca42363ac [flang] Add support for -f[no-]verbose-asm (#130788)
143bf95d41f4 [hwasan] Don't check code model if there are no globals (#131152)
b003face11fa [flang][OpenMP] Add `OutlineableOpenMPOpInterface` to `omp.teams` (#131109)
7661526fdf74 [X86] combineConcatVectorOps - extend PSHUFD/LW/HW handling to support 512-bit types
57e22f515e7a [X86] Add tests showing failure to concat matching VPSHUFLW/HW ymm shuffles.
85318bae285e [MachineLateInstrsCleanup] Handle multiple kills for a preceding definition. (#119132)
28ffa7f6a4d6 [flang][OpenMP] Fix missing missing inode issue (#130798)
237a9108190d [MLIR][OpenMP] Remove the ReductionClauseInterface, NFC (#130978)
c3c97eab1274 PeepholeOpt: Do not skip reg_sequence sources with subregs (#125667)
6ff33edf4d59 [MLIR][OpenMP] Minor improvements to BlockArgOpenMPOpInterface, NFC (#130789)
e3c80d4496d1 AMDGPU: Fix broken negative test from ancient times (#131106)
7a5e4f540580 [clang][NFCI] Fix getGridValues for unsupported targets (#131023)
739d1b4eb815 [Comgr] Add -nogpuinc to tests (#1062) (#1123)
5d5e706691c6 [VPlan] Restrict hoisting of broadcast operations using VPDominatorTree (#117138)
ffe202ca0053 Revert "[LV] Limits the splat operations be hoisted must not be defined by a recipe. (#117138)"
d4baf611c1cc [X86] combineConcatVectorOps - add outstanding TODOs for missing op concatenation cases. NFC.
0aa5ba43a0d1 [mlir] Fix DistinctAttributeUniquer deleting attribute storage when crash reproduction is enabled (#128566)
c26ec7ea2add [llvm][lit] fix writing results to --time-trace-output file (#130845)
b990c815029c Partial fix of BUILD.bazel after 598e882. (#131140)
b67379c35be7 [llvm-diff] Add colorful output to diff (#131012)
d1c4befe7ad2 [AMDGPU] Auto generated check lines for two tests (#1126)
7e9802f348e3 [Headers][NFC] Steps to allow sharing code between gpu intrin.h headers (#131134)
94c8fa61d697 AMDGPU: Replace some test i32 undef uses with poison (#131092)
024df9c9dc16 AMDGPU: Replace some float undef test uses with poison (#131090)
d507b3d7ce34 [X86] combineGatherScatter - pull out repeated index value type. NFC.
dff22a0c1173 [X86] combineConcatVectorOps - convert X86ISD::BLENDI concatenation to use combineConcatVectorOps recursion (#131121)
a413ef83a16b [lldb] Use Function::GetAddressRange*s* in "frame diagnose" (#130949)
d3255474be3e Reapply "[Offload][AMDGPU] LLVM_ENABLE_RUNTIMES=flang-rt for amdgpu-offload-*" (#130274)
0735537ec895 [Flang-RT] Append to library subdir
da69147a1619 [InferAttrs] Refine attributes for a few libc routines
1eb5588457c3 AMDGPU: Replace i16 undefs with poison in tests (#131084)
f613bc57c638 AMDGPU: Replace half undef uses with poison in tests (#131083)
b910610eeaf1 [libc][math] Fix Sollya command (#131091)
02fae68a45fd [mlir][vector] VectorLinearize: `ub.poison` support (#128612)
95d28fe503cc [Premerge] Add flang-rt (#128678)
786e70ff104f [mlir] Change `TypeOrValueSemanticsContainer` base from `TypeConstraint` to `Type` (#129433)
13261e856aa5 [Flang-RT][NFC] Fix comment
46739be7bcfc [GlobalISel] Preserve original flags of output instructions in matchtable (#130937)
55b806c2afb1 [MLIR][Affine] Fix affine data copy generation copy placement for missing memref definition check (#130750)
a5a162cd713d [SDAG] Pass pointer type to libcall expansion for SoftenFloatRes stack slots (#130647)
0c5eb4d68b09 [flang] Use precompiled parsing headers (#130600)
481a55a3d964 [rtsan][Apple] Add interceptor for _os_nospin_lock_lock (#131034)
abdbaff5441e [DWARFLinker] Adjust DW_AT_LLVM_stmt_sequence for rewritten line tables (#128953)
139add531a55 [InferAttrs] Mark floating-point libcalls as `errno`-writing
d0188ebcc206 [flang][OpenMP]Add symbls omp_in, omp_out and omp_priv in DECLARE RED… (#129908)
459b4e3fe108 Reland "[AMDGPU] Remove s_delay_alu for VALU->SGPR->SALU (#127212)" (#131111)
1282878c52cb [CI] Fix bad timestamps being reported (#130941)
d77ef140e66d [LoongArch] fix vec-trunc.ll test error (#131104)
de895751d2a2 [CaptureTracking][AA] Only consider provenance captures (#130777)
55b480ec3c87 [SROA] Allow load-only promotion with read-only captures (#130735)
59fd2878fce4 [mlir][memref] Clean up `load`/`store` documentation (#130569)
cb28ec6cccf9 [Sema] Instantiate destructors for initialized members (#128866)
08de320aa2f4 [lldb] Remove Function::GetAddressRange usage from the gui (#130991)
5952972c9164 [CodeGen][NPM] Port BranchFolder to NPM (#128858)
43ab4228d00a [LoongArch] Pre-commit test for vector trunc (#131082)
4f7ea8928d49 merge main into amd-staging
bd748b33958f [RISCV] Add implicit operand {VL, VTYPE} in RISCVInsertVSETVLI when u… (#130733)
6345b009c3e5 [C++20] [Modules] Add mangling number for lambda in non-internal module unit context
9e91725fd4d4 AMDGPU: Replace some undef uses in test metadata with poison (#131052)
26ae98c4d73d AMDGPU: Replace undef global initializers in tests with poison (#131051)
21cef8aa1c95 [MLIR][NVVM] Add support for tcgen05.{ld, st} (#130728)
7811075b6f66 AMDGPU: Replace ptr undef in tests with poison (#131050)
f81531c38be0 Move VE-specific MCSymbolRefExpr::VariantKind to VEMCExpr
331250c6fac8 AMDGPU: Replace ptr addrspace(3) undef in tests with poison (#131049)
34f4ee7c0d98 [C++20] [Modules] Merge codes to decide if we should generate decl
1d1f2e8c5be3 [M68k] Implement TLI::convertSelectOfConstantsToMath
6abe19ac587d [clang] Predefine `_CRT_USE_BUILTIN_OFFSETOF` in MS-compatible modes (#127568)
5d5f16204f17 Move PowerPC-specific MCSymbolRefExpr::VariantKind to PPCMCExpr
95e186cadfc8 Reland "DAG: Preserve range metadata when load is narrowed" (#128144) (#130609)
73e12de062c1 [HLSL] Implement explicit layout for default constant buffer ($Globals) (#128991)
bc4b2c74fe4e [X86][APX] Add NF instructions to convertToThreeAddress functions (#130969)
642a4763dfaa [PowerPC] Rename PPCMCExpr's VK_PPC_ to VK_. NFC
fe0d3e376496 [Sema] Diagnose by-value copy constructors in template instantiations (#130866)
646a6e7f1088 [PowerPC] Simplify PPCMCExpr::printImpl
08a3c532255d [RegAlloc] Scale the spill weight by target factor (#113675)
3438dfc7ff88 [mlir][tensor] Fix bufferization interface for 'tensor.reshape' (#128590)
376e3b62cd36 [TableGen] Add `!match` operator to do regex matching (#130759)
f291ec692e88 [libc] Use proxy header in the `locale` implementation. (#130982)
79a5974294eb [gn build] Port 4dcba5e08dc9
4dcba5e08dc9 [alpha.webkit.ForwardDeclChecker] Add a new WebKit checker for forward declarations (#130554)
62e37a8a0ae8 [RISCV][Disassembler] Use a table to store all the decoder tables and their associated features. NFC (#130883)
853849a984dd [ORC] Remove some unnecessary namespace qualifications. NFCI.
0a78bd67b311 AMDGPU: Make frexp_exp and frexp_mant intrinsics propagate poison (#130915)
4d8070e95602 [C++20] [Modules] Don't add decls from other units to undefinedButUsed set
d8f17b3de1ca AMDGPU: Make sqrt and rsq intrinsics propagate poison (#130914)
95ab95fd10b2 AMDGPU: Make rcp intrinsic propagate poison (#130913)
eea7d32bd262 [MC] Move fixSymbolsInTLSFixups to ELFObjectWriter
e4a0fd5416a3 [gn build] Port be9ca85d64eb
be9ca85d64eb [alpha.webkit.webkit.RetainPtrCtorAdoptChecker] Add a new WebKit checker for correct use of RetainPtr, adoptNS, and adoptCF (#128679)
716c34b15060 merge main into amd-staging (#1163)
12fe5ae88cce AMDGPU: Replace ptr addrspace(8) undef uses with poison (#130904)
27d83184c4ca [alpha.webkit.UncountedCallArgsChecker] Treat an explicit construction of Ref from a Ref return value safe. (#130911)
e1e44dfcad00 AMDGPU: Replace <8 x i32> undef uses in tests with poison (#130903)
c182f4042ada AMDGPU: Replace <4 x i32> undef uses in tests with poison (#130902)
6705d812b856 AMDGPU: Replace ptr addrspace(1) undefs with poison (#130900)
44ff94e99e03 [Diagnostics] Return rvalue reference from temporary argument (#127400)
cb73271be12a [AMDGPU][CodeGen] use vt in VGPRimm pattern (#131016)
f4fc2d731c1b [flang][OpenMP] Map ByRef if size/alignment exceed that of a pointer (#130832)
c54299170395 [flang-rt] Fixed HAVE_LDBL_MANT_DIG_113 detection. (#131010)
726ffd361cd9 [NFC][Cloning] Replace IdentityMD set with a predicate in ValueMapper (#129147)
98c279a3ed20 [lldb] Add missing optional conversion
1cfca53b9f2e [llvm][Timer] Don't print timers in TimerGroup when all Timers are removed (#131026)
c14e459ef8bd [Clang] add additional tests for -Wshift-bool (#130339)
ecf4d995f689 [mlir][linalg][elementwise] Fold transpose into new elementwise (#130207)
0639201ce8b3 merge main into amd-staging
be0215d7456f [CIR] Add transform test for cir-flatten-cfg (#130861)
62994c329123 [VPlan] Also introduce explicit broadcasts for values from entry VPBB.
9c65e6ac115a [HEXAGON] Add support to lower "FREEZE a half(f16)" instruction on Hexagon and fix the isel-buildvector-v2f16.ll assertion (#130977)
202137dbead8 [SLP]Fix a crash on matching gather operands of phi nodes in loops
64b94105d5c7 [CIR] Upstream support for emitting ignored statements (#130869)
202d9a7b51f3 Amd/dev/rlieberm/reland flang rt (#1159)
cba9dc6e9d04 [libc][nfc] Use common implementation of read_first_lane_u64 (#131027)
c476a4a907a5 [Clang][OpenCL] Fix Missing `-fdeclare-opencl-builtins` When Using `--save-temps` (#131017)
c8047c6dbcb5 [AMDGPU][True16][CodeGen] update test for buildbot failure (#131028)
15e6bb622417 [ORC] Rename wrapper arguments to match conventions. NFCI.
9d7e1d92dbef [AMDGPU][True16] added Pre-RA hint to improve copy elimination (#103366)
5929de8c7731 [gn build] Port 7790d69cce04
7790d69cce04 [lldb-dap] Refactoring IOStream into Transport handler. (#130026)
4d79e9892c48 [Support] Avoid repeated hash lookups (NFC) (#130891)
aa008e00085a Revert "[AMDGPU] Remove s_delay_alu for VALU->SGPR->SALU (#127212)"
598e882ee88a [libc] Template the writing mode for the writer class (#111559)
3aa96f52cff3 [OpenMP] [Taskgraph] Differentiating task ids from the taskgraph and from the debugger (#130660)
ab53e1c8e523 [C2y] Claim conformance to WG14 N3363 (#130980)
ed5d9c3abcbe merge main into amd-staging (#1155)
7f415e444197 [clang][SYCL] Disable float128 device mode diagnostic (#128513)
e27fe2e07c04 [OpenACC][NFC] Add the _Pragma spelling to a bunch of the tests
96be34e595bb [SandboxVec][SeedCollector][NFC] Fix typo in banner
c44c9051747e [mlir][tosa] Add error if verification to pooling operators (#130052)
f62e168d3f1d [lldb-dap] Validate server mode support prior to invoking lldb-dap. (#130855)
caf301891a10 Add unretained call args checker (#130901)
378739f18208 [SystemZ] Move disabling of arg verification to before isFullyInternal(). (#130693)
e619073f0302 [SandboxVec][NullPass][NFC] Add missing banner comment at top of file
8b093e5e1752 [webkit.UncountedLambdaCapturesChecker] Recognize std::move(protectedThis) (#130925)
184f9449572b Implement the `fmod` intrinsic (#130320)
78c9fa3a380e [lldb] Implement ANSI & Unicode aware string stripping & padding (#130878)
6030936ec82a [NVPTX] Fix bug in sign of bfe folding (#130862)
fc127ff53d0c [mlir] Extract RHS rows once when lowering vector.contract to dot (#130130)
40d6784ac441 [flang] preemptive sentinel for flang-rt if needed
1a626e63b5a0 [flang] Fix deprecation warning
e1ac57d53a18 [MemProf] Extend CallSite information to include potential callees. (#130441)
4518780c3cff [PowerPC] Add intrinsics and tests for basic Dense Math enablement instructions (#129913)
ea2e66aa8b6e [LLVM][ConstantFold] Undefined values are not constant (#130713)
dd181af9506d [X86] Add isMaskableNode helper to determine if a node could potentially fold into a AVX512 predicated instruction. NFC.
daecde2f84b8 [Flang] Fix libquadmath in non-LLVM_ENABLE_RUNTIMES build.
45c6d0dd3a96 [Flang-RT] Environment introspection for quadmath.h (#130411)
f137c3d592e9 [TargetRegistry] Accept Triple in createTargetMachine() (NFC) (#130940)
71582c6667a6 [AMDGPU] Remove s_delay_alu for VALU->SGPR->SALU (#127212)
ec941a4a0451 [NVPTX] Legalize ctpop and ctlz in operation legalization (#130668)
48b19912ca58 [NVPTX][test] Use 'not' to switch an XFAIL test to PASS (#130839)
f4d599cda90a [Support] Do not remove lock file on failure (#130834)
2620742bf61e [Utils] Avoid repeated hash lookups (NFC) (#130892)
a3b0189f6daa [Transforms] Avoid repeated hash lookups (NFC) (#130890)
db3fdbc84beb [CodeGen] Avoid repeated hash lookups (NFC) (#130889)
0359677695a7 [Driver] Avoid repeated hash lookups (NFC) (#130888)
e5aac528fede [AST] Avoid repeated hash lookups (NFC) (#130887)
bdbe8fa1f3dc [flang] Align `-x` language modes with `gfortran` (#130268)
1db978cd7813 [DirectX] Remove DXILResourceMDAnalysis (#130323)
665299eb3e7a [mlir][Transforms] Add a utility method to move value definitions. (#130874)
e11ede5e90ee Revert "[MS][clang] Add support for vector deleting destructors (#126240)"
90a8322399c7 [RISCV] Add an error that Xqccmp, Xqciac, and Xqcicm are not compatible with C+D or Zcd. (#130816)
d71b3debc966 [X86][GISel] Use Register and MCRegister. NFC (#130907)
7a25c725abc5 [X86] Use Register in X86SpeculativeLoadHardening.cpp. NFC (#130905)
5c02e74d21fd Fix use of CXXThisScopeRAII
10085390c6d4 [SLP]Reduce number of alternate instruction, where possible
9820248e0aaa AddressSanitizer: Add use-after-scope to pass options (#130924)
982527eef0f2 [flang] Use saturated intrinsics for floating point to integer conversions (#130686)
8be1d1235d58 [X86] combineX86ShufflesRecursively/combineX86ShuffleChain - pass down SDLoc instead of relying on Root node every recursion. NFC.
c56514f75b07 [X86] mfence.ll - remove dead X32 check prefix
07d86d25c9bc [ctxprof] Flat profile collection (#130655)
a502c656881c [BuildLibCalls] Add helper for setting memory effects (NFC)
190063464e12 [MC] Speed up checkFeatures() (NFCI) (#130936)
cbeae3e117b8 [Flang] Fix libquadmath in non-LLVM_ENABLE_RUNTIMES build.
5da9044c4084 [MemCpyOpt] Fix clobber check in fca2memcpy optimization
72ee9b82151f [NFC][analyzer] Rename `CheckerBase::getCheckerName` to `getName` (#130953)
2c6620f4f373 merge main into amd-staging
b76e396990ef AMDGPU: Replace tests using undef in shufflevector with poison (#130899)
2fbddfbdc014 [X86] combineConcatVectorOps - remove unused DAGCombinerInfo argument. NFC. (#130951)
c6e88b218942 Proper BUILD.bazel fix due to c07e1e3 (#130962)
1d0dd76eec05 AMDGPU: Replace insertelement undef with poison in cases with manual updates (#130898)
da42b2f67da6 AMDGPU: Replace insertelement poison with insertelement undef (#130896)
ab557afa40af [libc][nfc] Include instantiations of gpuintrin.h in IR test case (#130956)
2c8e154ce8e8 [gn build] Port f10a8706a144
6af9fd9430d5 [gn build] Port 76d5a79bed00
1e58bdcc73c5 [gn build] Port 75f76d482cc2
cd043e4fbe61 [Docs] Add a link to Clang Area Team members
de34213a67c2 [gn] port ab6f470675ed
6717a38bb5fd [gn] port 3a7a9c928671 better
2e921452ca09 [Clang] [OpenMP] Fixed wrong type usage in Xteam Reduction codegen. (#1141)
72281311adc1 merge main into amd-staging (#1148)
73e23f899f2a [AMDGPU] Change placeholder from `undef` to `poison` (#130858)
15136f1c09d9 [AMDGPU] Change placeholder from `undef` to `poison` (#130853)
5073b5fdfaf9 [CVP] Infer `nuw`/`nsw` flags for TruncInst (#130504)
059ada405c1e [PreISelintrinsicLowering] getTypeSizeInBits/8 => getTypeAllocSize in memset.pattern lowering
f93cd423bdd2 [Comgr] Turn DEVICE_LIBS_ID into an unsigned char (#1127)
1d89d7d5d76e [Docs] Explain how to propose an extension in Clang (#130803)
6bf0c4648eea reapply [llvm] add support for mustache templating language (#130876)
cf68c9378b0c [Flang][OpenMP] Move declare mapper sym creation outside loop, NFC (#130794)
032f83b743b7 [MLIR][OpenMP] Enable BlockArgOpenMPOpInterface accessing operands (#130769)
c851ee38ad45 [flang][OpenMP] catch namelist access through equivalence (#130804)
0165ca3a0938 BUILD.bazel fix due to c07e1e3
76cf895717e9 Revert "[HLSL] error on out of bounds vector accesses (#128952)"
30fa7a231def [NFC][Cloning] Make DifferentModule case more obvious in CollectDebugInfoForCloning (#129146)
77b55c76aceb [llvm][docs] Add notes on upstreaming code from downstream projects (#129743)
014bf63c2e7f AMDGPU/GlobalISel: Temporal divergence lowering i1 (#124299)
c07e1e390ccb AMDGPU/GlobalISel: Temporal divergence lowering (non i1) (#124298)
51fa25b2ee49 [Comgr][Cache] Fix broken test: spirv-translator-cached.cl
553da9634dc4 AMDGPU/GlobalISel: Update divergence lowering tests (#128702)
fef0b8a0ba58 [lldb] Let languages see all SymbolContexts at once when filtering breakpoints (#129937)
5ec884e5d8a1 Revert "[SLP]Reduce number of alternate instruction, where possible"
3ad810ea9a97 AMDGPU/GlobalISel: Disable LCSSA pass (#124297)
9f617161aa8d [CodeGen][NPM] Port PatchableFunction to NPM (#129866)
ad704ff62b4a [X86][NF] Switch the order of Inst and &Target.getInstruction(NewRec) (#130739)
fe134d47c37c merge main into amd-staging
65ade6d2ebbe [AMDGPU] Merge consecutive wait_alu instruction (#128916)
adb44ed2b864 [AArch64] Add -cost-kind=all coverage for insert-extract.ll and shuffle-load.ll. NFC
6f89c1ff6b88 [AArch64] Remove Kyro run lines from insert-extract.ll. NFC
5a0a2f8239d3 AMDGPU: Replace undef with poison in tests using insertvalue (#130895)
5daba3dfd927 [AMDGPU][NFC] Format GCNCreateVOPD.cpp (#130548)
c22c5643db68 [AMDGPU][NPM] Port SIMemoryLegalizer to NPM (#130060)
a6089a949fef [AMDGPU] Ignore RegMask operands when folding operands to SALU insts (#130813)
15e335f04fba [flang] also set llvm ABI argument attributes on direct calls (#130736)
c9563a422cea [Clang][NFC] Remove CallExpr::CreateTemporary (#130919)
369c0a7483a4 [clang-repl] Fix target creation in Wasm.cpp (#130909)
525d412cae53 [AMDGPU] Fix typing error introduce in promote alloca change
7decd046260d [Clang] Add __builtin_elementwise_exp10 in the same fashion as exp/exp2 (#130746)
418e07b7e679 [mlir][Tensor] Check for out-of-bounds slice in `insert/extract_slice` verifier (#130487)
c86d88484709 AMDGPU: Update LIT tests for i1-to-bf16 conversions (NFC) (#130916)
528a6979543f merge main into amd-staging (#1140)
90a08fb4b7e7 [LLD][COFF] Update nodefaultlibs after updating search paths (#128813)
d921bf233c4f [AMDGPU] Extend promotion of alloca to vectors (#127973)
57a90883ca54 [CodeGen][NPM] Port DetectDeadLanes to NPM (#130567)
8d7cb5d05b12 [HLSL][NFC] Update resource metadata tests to not use obsolete annotations (#130222)
bde140c47fe0 AMDGPU: Remove a FileCheck from an XFAILed test
5f20f9a01261 [clang][AST] Remove HasFirstArg assertion in CallExpr::getBeginLoc() (#130725)
606e9fa44455 [WindowsDriver] Always consider `WinSdkVersion` (#130377)
7a66a26658f4 [llvm-objcopy,ELF] --discard-locals/--discard-all: allow and keep symbols referenced by relocations
1585db458f04 [IR] Optimize CFI in `writeCombinedGlobalValueSummary` (#130382)
bf2d1c46072a AMDGPU: Disable machine verifier in failing test
d898761ca24a [RISCV] FeatureVendorXwchc should imply FeatureStdExtZca. (#130817)
725726171240 [NVPTX] Fix generic address in st.bulk intrinsic (#130740)
adae90ee35ca AMDGPU: Add baseline test for coalescing vgpr to agpr subreg inserts (#130877)
e427f0694ed0 [ADT] Use adl_begin in make_first_range and make_second_range (#130521)
9ef7287d4252 [profile] runtime counter relocation needed on all windows targets (#127858)
c02b0c99f0c1 [OpenMP][Offload][AMDGPU] Add envar for setting CU multiplier
ab22f652a4df [libc] implement `strings/str{n}casecmp_l` (#130407)
c12761858c72 [Clang] Fix the printout of CXXParenListInitExpr involving default arguments (#130731)
30fdeec0f8f9 [ADT] Use `adl_being`/`end` in `map_range` (#130508)
6981f7e92a05 [mlir] account for explicit affine.parallel in parallelization (#130812)
554347ba45b5 Revert "[llvm] add support for mustache templating language (#105893)" (#130873)
7ae75851b2e1 [X86][APX] Support peephole optimization with CCMP instruction (#129994)
9415b7d97fc3 [Support] Fix -Wpessimizing-move in Mustache.cpp (NFC)
371ea32bafc7 breaks flang build: Revert [Flang-RT] Environment introspection for quadmath.h (#130411)
8f05f2536016 reapply [llvm] add support for mustache templating language (#130732)
b6f502ddb0d1 AMDGPU: Implement i1 to bfloat conversion (#130831)
da708814beee [RISCV] Move DecoderNamespace in RISCVInstrInfoXCV.td to the Instruction defs. NFC (#130800)
da478df56358 [X86] Avoid repeated hash lookups (NFC) (#130710)
72aec1dfcae4 [IPO] Avoid repeated hash lookups (NFC) (#130708)
bc5153b987a6 [Clang] [OpenMP] Fixed wrong type usage in Xteam Reduction codegen.
3692fb6ba58a [mlir][math] add benefit arg to populate math approximations/expansions (#130782)
d547005f6c2b [flang-rt] Enable -funwind-tables so that backtrace works. (#130848)
050d1cbf865c merge main into amd-staging
76d5a79bed00 [ORC] Drop EHFrameRegistrar, register eh-frames with AllocActions (#130719)
634e25319e0e [mlir] Add special case for 0-D tensor when fusing expand from collapse (#130838)
701148f05a7b [CIR] Upstream initial support for CIR flattening (#130648)
6dbb5319ce3e [flang-rt] Set HAVE_LDBL_MANT_DIG_113. (#130836)
0e4ba47ca827 [clang-tidy] support to detect conversion in `make_optional` for `bugprone-optional-value-conversion` (#130417)
4d6ca116221a [AstMatcher]`templateArgumentCountIs` support `FunctionDecl` (#130416)
009dfb435f3e Reland "[LLD][ELF] Don't spill to same memory region" (#130851)
89fa592b68c7 [mlir][tosa] Update Matmul description to align with spec (#130835)
34a3c2302ba5 [AArch64][SVE] Change placeholder from `undef` to `poison` (#130519)
dc23234a6613 [VPlan] Remove dead code in VPWidenPHIRecipe::print (NFC).
8c97ddff5318 [mlir][DataLayout] Add a default memory space entry to the data layout. (#127416)
dad0a4e88636 [ADT] Use `adl_being`/`adl_end` in `make_early_inc_range` (#130518)
486faec155bf Amd/dev/agozillo/rebase reland has addr (#1128)
8132c4f55474 [VPlan] Also introduce broadcasts for live-ins used in vec preheader.
74eba972ca38 [flang] Definitions of fir.pack/unpack_array operations. (#130698)
d578148b7d52 [libc][math] Skip checking for exceptional values when LIBC_MATH_SKIP_ACCURATE_PASS is set. (#130811)
c337e2d250a9 [NFC][AMDGPU] Auto generate check lines for some disassembler tests (#130799)
3189402466d3 [NFC][AMDGPU] Auto generate check lines for some assembler tests (#130797)
f9568e8d23b7 [HLSL] Make memory representation of boolean vectors in HLSL, vectors of i32. Add support for boolean swizzling. (#123977)
dafb566710cd [Support] Return `LockFileManager` errors right away (#130627)
7573ee17813b Revert "[alpha.webkit.UnretainedCallArgsChecker] Add a checker for NS or CF type call arguments." (#130828)
d84755b10d09 [docs] Fix broken formatting for link in HowToCrossCompileLLVM
10a6a349d66a [libc][bazel] Create libc_release_library for release configurations. (#130694)
2f403ee48714 [clang][modules] NFC: Remove unused function parameter
69f59285ea27 [X86] combineConcatVectorOps - convert X86ISD::PCMPEQ/PCMPGT concatenation to use combineConcatVectorOps recursion instead of IsConcatFree (#130814)
9e64fc6fb7d9 [alpha.webkit.UnretainedCallArgsChecker] Add a checker for NS or CF type call arguments. (#130729)
c2ed840ed94d Revert "[LLD][ELF] Don't spill to same memory region" (#130815)
1cb14078658f [HLSL] Add bounds checks for the hlsl vector arguments and return types (#130724)
456fa47bab47 [C2y] Claim conformance to WG14 N3505
381599f1fe97 [ELF] Allow KEEP within OVERLAY (#130661)
14176d10842d [C2y] Remove WG14 N3459 from the C status page
7de895ff1146 [SLP]Reduce number of alternate instruction, where possible
b81fc28f8b4d [Comgr][NFC] Add space
ba11e1e5222c [docs] Rewrite HowToCrossCompileLLVM (#129451)
5686786c550c [BPF] Fix BitCast Assertion with NonZero AddrSpace (#130722)
3fc3e7b378ea [Comgr] Re-populate DataAction options for second compilation
313e2ef93eb2 [libc] Mark fixed point type generic macros as complete (#130805)
9f2bd97a5dcb [X86] combineConcatVectorOps - convert X86ISD::PACKSS/US concatenation to use combineConcatVectorOps recursion (#130575)
c4280db3a0fe [X86] Add getBLENDIBlendMask helper to extract X86ISD::BLENDI blend mask. NFC.
5f21ee20f84a [LLD][ELF] Don't spill to same memory region (#129795)
f696ba95e861 merge main to amd-staging (#1118)
eaca60d1a9a5 [clang] Hide the `DiagnosticOptions` pointer from `CompilerInvocation` (#106274)
65016475084f [StaticAnalyzer] Relax the pre-condition of 'setsockopt' (#130683)
f10a8706a144 [analyzer] Sink false [[assume]] execution paths (#130418)
d22d14375d44 [clang-tidy] support different precisions (#130540)
517c6778ead6 [clang-tidy] Add check on constexpr & static values in modernize-use-default-member-init (#129425)
4f60f45130c6 [llvm] replace static_assert with std::enable_if_t in ilist_node_impl (#127722)
f1598367b661 [lldb-dap] Adding logging helpers. (#130653)
1324dfe3e177 [AIX] Add -pthread to build on AIX (#129108)
2bf7018bb03e [RISCV] Move let statement for hasSideEffects, mayLoad, mayStore into BranchCC_rri. NFC (#130721)
90c11ad46f42 [ELF] Introduce ReportPolicy to handle -z *-report options. NFC
34647667aa02 [RISCV] Sink hasSideEffects, mayLoad, mayStore from defs to classes in RISCVInstrInfoXCV.td. NFC (#130714)
146ef7a5f44a [TableGen] Remove unnecessary const_cast and use range-based for loops. NFC (#130717)
83ec179fc8b4 [Clang][NFC] Rename and update_cc_test_checks over strictfp-elementwise-builtins.cpp (#130747)
74c3df6cb554 [Comgr] Turn DEVICE_LIBS_ID into an unsigned char
0e13a2df117f Reland/fix conflicts for PR: d67947162f4b
ae985267d0a1 [RISCV] Update to Xqciint v0.4 (#130219)
f90aa4189787 [Support] Remove output file checks from `LockFileManager` (#130395)
93d41d814816 [LV] Use ElementCount::isKnownLT to factor code (NFC) (#130596)
1fe463182cea [hexagon] Prevent alignment search beyond a label (#130631)
d8d2e0779a40 Even more BUILD.bazel fixes for commit 205c532. (#130784)
a71c9d843862 [NFC][analyzer] Remove CheckerNameRef::getName() (#130780)
c542f425796d [AArch64] Update cost test to use -cost-kind=all. NFC
f3e55944a946 [mlir][ODS] Switch declarative rewrite rules to properties structs (#124876)
b334321678d4 [X86] Prefer `lock or` over mfence (#106555)
222b99d3aac5 [AMDGPU][True16][CodeGen] update waitcnt for true16 (#128927)
7129205816ca [LoopVectorize] Move checking for OptForSize into the cost model (NFC) (#130752)
fb397ab1e5f3 Reland "[clang] Lower modf builtin using `llvm.modf` intrinsic" (#130761)
278087cc86db [Comgr] Add -nogpuinc to tests (#1062)
7c77a4655ca5 [X86] combineConcatVectorOps - convert ISD::VECTOR_SHUFFLE concatenation to use combineConcatVectorOps recursion (#130610)
f33dca41a3a0 [llvm-rtdyld] Avoid repeated hash lookups (NFC) (#130711)
3339632e9c68 [Utils] Avoid repeated hash lookups (NFC) (#130709)
8c2714e44802 [ExecutionEngine] Avoid repeated hash lookups (NFC) (#130707)
21f1ef330f22 [AMDGPU] Avoid repeated hash lookups (NFC) (#130706)
d48a36f5833a [ms] [llvm-ml] Allow optional parenthesized arguments for macros (#129905)
35…
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
This patch updates fir.coordinate_op to carry the field index as attributes instead of relying on getting it from the fir.field_index operations defining its operands.
The rational is that FIR currently has a few operations that require DAGs to be preserved in order to be able to do code generation. This is the case of fir.coordinate_op, which requires its fir.field operand producer to be visible.
This make IR transformation harder/brittle, so I want to update FIR to get rid if this.
Codegen/printer/parser of fir.coordinate_of and many tests need to be updated after this change.
After this patch similar changes should be done to make FIR more robust: