This commit restructures how TypeID is implemented to ideally avoid
the current problems related to shared libraries. This is done by changing
the "implicit" fallback path to use the name of the type, instead of using
a static template variable (which breaks shared libraries). The major downside to this
is that it adds some additional initialization costs for the implicit path. Given the
use of type names for uniqueness in the fallback, we also no longer allow types
defined in anonymous namespaces to have an implicit TypeID. To simplify defining
an ID for these classes, a new `MLIR_DEFINE_EXPLICIT_INTERNAL_INLINE_TYPE_ID` macro
was added to allow for explicitly defining a TypeID directly on an internal class.
To help identify when types are using the fallback, `-debug-only=typeid` can be
used to log which types are using implicit ids.
This change generally only requires changes to the test passes, which are all defined
in anonymous namespaces, and thus can't use the fallback any longer.
Differential Revision: https://reviews.llvm.org/D122775
Apply scale should be optionally disabled when lowering via TosaToStandard.
In most cases it should persist until the lowering to specific backend.
Reviewed By: jpienaar
Differential Revision: https://reviews.llvm.org/D122948
Adds the ability to create external passes using the C-API. This allows passes
to be written in C or languages that use the C-bindings.
Differential Revision: https://reviews.llvm.org/D121866
This patch changes the implementation of SetCoalescer to use PresburgerSpace
instead of reimplementing parts of PresburgerSpace.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D122984
Add integer-exact checks for inequalities being separate and redundant in LexSimplex.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122921
This significantly simplifies the boilerplate necessary for passes
to define nested pass pipelines.
Differential Revision: https://reviews.llvm.org/D122880
ListOption currently uses llvm:🆑:list under the hood, but the usages
of ListOption are generally a tad different from llvm:🆑:list. This
commit codifies this by making ListOption implicitly comma separated,
and removes the explicit flag set for all of the current list options.
The new parsing for comma separation of ListOption also adds in support
for skipping over delimited sub-ranges (i.e. {}, [], (), "", ''). This
more easily supports nested options that use those as part of the
format, and this constraint (balanced delimiters) is already codified
in the syntax of pass pipelines.
See https://discourse.llvm.org/t/list-of-lists-pass-option/5950 for
related discussion
Differential Revision: https://reviews.llvm.org/D122879
Prior to this change there were a number of places where the allocation and deallocation of SparseTensorCOO objects were not cleanly paired, leading to inconsistencies regarding whether each function released its tensor/coo arguments or not, as well as making it easy to run afoul of memory leaks, use-after-free, or double-free errors. This change cleans up the codegen vs runtime boundary to resolve those issues. Now, the only time the runtime library frees an object is either (a) because it's a function explicitly designed to do so, or (b) because the allocated object is entirely local to the function and would be a memory leak if not released. Thus, now the codegen takes complete responsibility for releasing any objects it caused to be allocated.
Reviewed By: aartbik
Differential Revision: https://reviews.llvm.org/D122435
This patch modifies the name "integerRelations" and "relation" to refer to the
disjuncts in PresburgerRelation to "disjunct(s)". This is done to be
consistent with the rest of the interface.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D122892
Previously, when an input set had a duplicate division, the duplicates might
be removed by a call to mergeLocalIds due to being detected as being duplicate
for the first time. The subtraction implementation cannot handle existing
locals being removed, so this would lead to unexpected behaviour. Resolve this
by removing all the duplicates up front.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122826
LexSimplex cannot be made to support symbols for symbolic lexmin; this requires
a second class. In preparation for upstreaming support for symbolic lexmin,
keep the part of LexSimplex that are specific to non-symbolic lexmin in LexSimplex
and move the parts that are required to a common class LexSimplexBase for both to
inherit from.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122828
This patch modifies IntegerPolyhedron, IntegerRelation, PresburgerRelation,
PresburgerSet, PWMAFunction, constructors to take PresburgerSpace instead of
dimensions. This allows information present in PresburgerSpace to be carried
better and allows for a general interface.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D122842
This patch supports ordered clause specified without parameter in
worksharing-loop directive in the OpenMPIRBuilder and lowering MLIR to
LLVM IR.
Reviewed By: Meinersbur
Differential Revision: https://reviews.llvm.org/D114940
This reverts commit 59bbc7a085.
This exposes an issue breaking the contract of
`applyPatternsAndFoldGreedily` where we "converge" without applying
remaining patterns.
This avoids accidentally reversing the order of constants during successive
application, e.g. when running the canonicalizer. This helps reduce the number
of iterations, and also avoids unnecessary changes to input IR.
Fixes#51892
Differential Revision: https://reviews.llvm.org/D122692
Bubble up extract_slice above Linalg operation.
A sequence of operations
%0 = linalg.<op> ... arg0, arg1, ...
%1 = tensor.extract_slice %0 ...
can be replaced with
%0 = tensor.extract_slice %arg0
%1 = tensor.extract_slice %arg1
%2 = linalg.<op> ... %0, %1, ...
This results in the reduce computation of the linalg operation.
The implementation uses the tiling utility functions. One difference
from the tiling process is that we don't need to insert the checking
code for the out-of-bound accesses. The use of the slice itself
represents that the code writer is sure about the boundary condition.
To avoid adding the boundary condtion check code, `omitPartialTileCheck`
is introduced for the tiling utility functions.
Differential Revision: https://reviews.llvm.org/D122437
This patch fixes a bug in LinearTransform::applyTo where it did not carry the
IdKind information, and instead treated every id as IdKind::Domain.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D122823
Previously, when updating the outputs matrix, the local offset was not being considered.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122812
Infer a tighter MemRef type instead of always falling back to the most dynamic MemRef type. This is inefficient and caused op verification errors.
Differential Revision: https://reviews.llvm.org/D122649
* Complete rewrite of the verifier.
* CollapseShapeOp verifier will be updated in a subsequent commit.
* Update and expand op documentation.
* Add a new builder that infers the result type based on the source type, result shape and reassociation indices. In essence, only the result layout map is inferred.
Differential Revision: https://reviews.llvm.org/D122641
For example, we could do the following eliminations:
fold vector.shuffle V1, V2, [0, 1, 2, 3] : <4xi32>, <2xi32> -> V1
fold vector.shuffle V1, V2, [4, 5] : <4xi32>, <2xi32> -> V2
Differential Revision: https://reviews.llvm.org/D122706
This reverts commit 5630143af3. The
implementation in this commit was incorrect. Also, handling this representation
of equalities in the upcoming support for symbolic lexicographic minimization
makes that patch much more complex. It will be easier to review that without
this representaiton and then reintroduce the fixed column representation
later, hence the revert rather than a bug fix.
linalg.generic can also take scalars instead of tensors, which
tensor.cast doesn't support. We don't have an easy way to cast between
scalars and tensors so just keep the linalg.generic in those cases.
Differential Revision: https://reviews.llvm.org/D122575
We are using "enable-index-optimizations" and "indexOptimizations" as
names for an optimization that consists of using i32 for indices within
a vector. For instance, when building a vector comparison for mask
generation. The name is confusing and suggests a scope beyond these
vector indices. This change makes the function of the option explicit
in its name.
Differential Revision: https://reviews.llvm.org/D122415
A Block is optionally allocated & leaks in case of failed parse. Inline the
function and ensure Block gets freed unless parse is successful.
Differential Revision: https://reviews.llvm.org/D122112
The change of https://reviews.llvm.org/D121513#3411651
has caused a build error when building with clang:
/mnt/vss/_work/1/llvm-project/mlir/lib/Dialect/Tosa/IR/TosaOps.cpp:599:26: error: extra ';' outside of a function is incompatible with C++98 [-Werror,-Wc++98-compat-extra-semi]
ReduceFolder(ReduceAllOp);
Reviewed By: hpmorgan, Mogball
Differential Revision: https://reviews.llvm.org/D122599
Adds another pointer to the union in RegionRange to allow RegionRange to work on ArrayRef<Region *> (i.e. vectors of Region *).
Reviewed By: rriddle
Differential Revision: https://reviews.llvm.org/D122514
(This was a TODO from the initial patch).
The control-flow sink utility accepts a callback that is used to sink an operation into a region.
The `moveIntoRegion` is called on the same operation and region that return true for `shouldMoveIntoRegion`.
The callback must preserve the dominance of the operation within the region. In the default control-flow
sink implementation, this is moving the operation to the start of the entry block.
Reviewed By: mehdi_amini
Differential Revision: https://reviews.llvm.org/D122445
This has been on _Both for a couple of weeks. Flip usages in core with
intention to flip flag to _Prefixed in follow up. Needed to add a couple
of helper methods in AffineOps and Linalg to facilitate a pure flag flip
in follow up as some of these classes are used in templates and so
sensitive to Vector dialect changes.
Differential Revision: https://reviews.llvm.org/D122151
- Adds default implementations of `isDefinedOutsideOfLoop` and `moveOutOfLoop` since 99% of all implementations of these functions were identical
- `moveOutOfLoop` takes one operation and doesn't return anything anymore. 100% of all implementations of this function would always return `success` and uses would either respond with a pass failure or an `llvm_unreachable`.
This revision supports padding only a subset of the iteration dimensions via an additional padding-dimensions parameter. This control allows us to pad an operation in multiple steps. For example, one may want to pad only the output dimensions of a producer matmul fused into a consumer loop nest, before tiling and padding its reduction dimension.
Depends On D122309
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D122560
Pass the padding options using arrays instead of lambdas. In particular pass the padding value as string and use the argument parser to create the padding value. Arrays are a more natural choice that matches the current use cases and avoids converting arrays to lambdas.
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D122309
This patch adds the ReductionClauseInterface and also adds reduction
support for `omp.parallel` operation.
Reviewed By: kiranchandramohan
Differential Revision: https://reviews.llvm.org/D122402
This patch adds MLIR NVVM support for the various NVPTX `mma.sync`
operations. There are a number of possible data type, shape,
and other attribute combinations supported by the operation, so a
custom assebmly format is added and attributes are inferred where
possible.
Reviewed By: ThomasRaoux
Differential Revision: https://reviews.llvm.org/D122410
Use "enable-vla-vectorization=vla" to generate a vector length agnostic
loops during vectorization. This option works for vectorization strategy 2.
Differential Revision: https://reviews.llvm.org/D118379
The way vector.create_mask is currently lowered is
vector-length-dependent, and therefore incompatible with scalable vector
types. This patch adds an alternative lowering path for create_mask
operations that return a scalable vector mask.
Differential Revision: https://reviews.llvm.org/D118248
This patch is a cleanup patch that merges PresburgerLocalSpace and
PresburgerSpace. Asserting that there are no locals is shifted to the
users of PresburgerSpace themselves.
The reasoning for this patch is that PresburgerLocalSpace did not contribute
much and only introduced additional complexity as locals could still be present
in PresburgerSpace, just not writable. This could introduce problems if a
PresburgerSpace with locals was copied to PresburgerLocalSpace which expected
no locals in a PresburgerSpace.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D122353
This transformation allow to break up a reduction dimension in a
parallel and a reduction dimension. This is followed by a separate
reduction op. This allows to generate tree reduction which is beneficial
on target allowing to take advantage parallelism.
Differential Revision: https://reviews.llvm.org/D122045
This patch adds the nowait parameter to `createSingle` in
OpenMPIRBuilder and handling for IR generation from OpenMP Dialect.
Also added tests for the same.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D122371
Previously, only LinalgOps whose operands are defined by an ExtractSliceOp could be padded. The revision supports walking a use-def chain of LinalgOps to find an ExtractSliceOp.
Reviewed By: hanchung
Differential Revision: https://reviews.llvm.org/D122116
This revision introduces a heuristic to stop fusion for shape-only tensors. A shape-only tensor only defines the shape of the consumer computation while the data is not used. Pure producer consumer fusion thus shall not fuse the producer of a shape-only tensor. In particular, since the shape-only tensor will have other uses that actually consume the data.
The revision enables fusion for consumers that have two uses of the same tensor. One as input operand and one as shape-only output operand. In these cases, we want to fuse only the input operand and avoid output fusion via iteration argument.
Reviewed By: hanchung
Differential Revision: https://reviews.llvm.org/D120981
Create the AffineMinOp used to compute the padding width in canonical form and update the tests.
Reviewed By: springerm
Differential Revision: https://reviews.llvm.org/D122311
This patch adds translation from omp.single to LLVM IR.
Depends on D122288
Reviewed By: ftynse, kiranchandramohan
Differential Revision: https://reviews.llvm.org/D122297
Add a baseline implementation of support for local ids for `PWMAFunction::valueAt`. This can be made more efficient later if needed by handling locals with known div representations separately.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122144
In LexSimplex, instead of adding equalities as a pair of inequalities,
add them as a single row, move them into the basis, and keep them there.
There will always be a valid basis involving all non-redundant equalities. Such
equalities will then be ignored in some other operations, such as when looking
for pivot columns. This speeds them up a little bit.
More importantly, this is an important precursor patch to adding support for
symbolic integer lexmin, as this heuristic can sometimes make a big difference there.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122165
This simplifies many places where we just want to do something in a "transient context"
and return some value.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122172
It is often necessary to "rollback" IntegerRelations to an earlier state. Although providing full rollback support is non-trivial, we really only need to support the case where the only changes made are to append ids or append constraints, and then rollback these additions. This patch adds support to rollback in such situations by recording the number of ids and constraints of each kind and providing support to truncate the IntegerRelation to those counts by removing appended ids and constraints. This already simplifies subtraction a little bit and will also be useful in the implementation of symbolic integer lexmin.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122178
This provides a way to create an operation without manipulating
OperationState directly. This is useful for creating unregistered ops.
Reviewed By: rriddle, mehdi_amini
Differential Revision: https://reviews.llvm.org/D120787
Emitting at error at EOF will emit the diagnostic past the end of the file. When emitting an error during parsing at EOF, emit it at the previous character.
Reviewed By: jpienaar
Differential Revision: https://reviews.llvm.org/D122295
This fixes a bufferization issue with ops that are not supported by the buffer deallocation pass when `allow-return-allocs=0`.
Differential Revision: https://reviews.llvm.org/D122304
This patch adds omp.single according to Section 2.8.2 of OpenMP 5.0.
Also added tests for the same.
Reviewed By: peixin
Differential Revision: https://reviews.llvm.org/D122288
Co-authored-by: Kiran Kumar T P <kirankumar.tp@amd.com>
This is a convenience function for adding new divisions to the Simplex given the numerator and denominator.
This will be needed for symbolic integer lexmin support.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122159
This patch
- adds assembly format for `omp.wsloop` operation
- removes the `parseClauses` clauses as it is not required anymore
This is expected to be the final patch in a series of patches for replacing
parsers for clauses with `oilist`.
Reviewed By: Mogball
Differential Revision: https://reviews.llvm.org/D121367
I'm using "shape" to mean the compile-time object, where zeros indicate sizes which are compile-time dynamic; and using "sizes" to mean the run-time object, where zeros indicate a dimension with no coordinates (hence resulting in trivial storage). Because their semantics differ on zeros, it's important to keep them distinguished. Although we do not define separate C++ types to capture the distinction, we can at least use variable names to do so.
This is (tangential) work towards fixing: https://github.com/llvm/llvm-project/issues/51652
Depends On D122057
Reviewed By: aartbik
Differential Revision: https://reviews.llvm.org/D122058
This is work towards: https://github.com/llvm/llvm-project/issues/51652
This differential sets up the options and threads them through everywhere, but doesn't actually use them yet. The differential that finally makes use of them is D122061, which is the final differential in the chain that fixes bug 51652.
Reviewed By: aartbik
Differential Revision: https://reviews.llvm.org/D122054
This is the more logical place for the function to live. If/when we factor out a separate class for just the `Coordinates` themselves, then the definition should be moved to `Coordinates::lexOrder` (and `Element::lexOrder` would become a thin wrapper delegating to that function).
This is (tangentially) work towards fixing: https://github.com/llvm/llvm-project/issues/51652
Reviewed By: aartbik
Differential Revision: https://reviews.llvm.org/D122057
Previously, an UndoLogEntry was added by addRow but not by addZeroRow. So
calling directly into addZeroRow, as LexSimplex::addCut does, was not an
undoable operation. In the current usage of addCut this could never
lead to an incorrect result, and addZeroRow is protected, so it is not
currently possible to add a regression test for this. This bug needs to be
fixed for the symbolic integer lexmin algorithm.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D122162
The revision introduces a affine.min and affine.max canonicalization pattern that orders the result expressions. It flattens the result expressions to arrays of dimension and symbol coefficients plus one constant coefficient and rearranges them in lexicographic order.
Without the pattern, CSE will not eliminate two affine.min / affine.max operation if the results are ordered differently. For example, the operations
```
%1 = affine.min affine_map<(d0) -> (8, -d0 + 27)>(%arg4)
%2 = affine.min affine_map<(d0) -> (-d0 + 27, 8)>(%arg4)
```
doe not CSE. After applying the pattern, the two operations are equivalent
```
%1 = affine.min affine_map<(d0) -> (8, -d0 + 27)>(%arg4)
%2 = affine.min affine_map<(d0) -> (8, -d0 + 27)>(%arg4)
```
which enables CSE.
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D121819
Computing dropped unit-dims when all the unit dims are dropped, does
not need to check for strides being dropped.
This also enables canonicalization of reduced-rank subviews.
Reviewed By: gysit
Differential Revision: https://reviews.llvm.org/D121766
I am not sure about the meaning of Type in the name (was it meant be interpreted as Kind?), and given the importance and meaning of Type in the context of MLIR, its probably better to rename it. Given the comment in the source code, the suggestion in the GitHub issue and the final discussions in the review, this patch renames the OperandType to UnresolvedOperand.
Fixes https://github.com/llvm/llvm-project/issues/54446
Differential Revision: https://reviews.llvm.org/D122142
The current nested if merging has a bug. Specifically, consider the following code:
```
%r = scf.if %arg3 -> (i32) {
scf.if %arg1 {
"test.op"() : () -> ()
}
scf.yield %arg0 : i32
} else {
scf.yield %arg2 : i32
}
```
When the above gets merged, it will become:
```
%r = scf.if %arg3 && %arg1-> (i32) {
"test.op"() : () -> ()
scf.yield %arg0 : i32
} else {
scf.yield %arg2 : i32
}
```
However, this means that when only %arg3 is true, we will incorrectly return %arg2 instead
of %arg0. This change updates the behavior of the pass to only enable nested if merging where
the outer yield contains only values from the inner if, or values defined outside of the if.
In the case of the latter, they can turned into a select of only the outer if condition, thus
maintaining correctness.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D122108
When the current implementation merges two blocks that have operands defined outside of their block respectively, it will merge these by adding a block argument in the resulting merged block and adding successor arguments to the predecessors.
There is a special case where this is incorrect however: If one of predecessors terminator produce the operand, inserting the block argument and updating the predecessor would lead to the terminator using its own result as successor argument.
IR Example:
```
%0 = "test.producing_br"()[^bb1, ^bb2] {
operand_segment_sizes = dense<0> : vector<2 x i32>
} : () -> i32
^bb1:
"test.br"(%0)[^bb4] : (i32) -> ()
```
where `^bb1` is then merged with another block would lead to:
```
%0 = "test.producing_br"(%0)[^bb1, ^bb2]
```
This patch fixes that issue during clustering by making sure that if the operand is from an outside block, that it is not produced by the terminator of a predecessor.
Differential Revision: https://reviews.llvm.org/D121988
This patch adds translation from `omp.atomic.capture` to LLVM IR. Also
added tests for the same.
Depends on D121546
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D121554
This patch improves the representation size of individual
`IntegerRelation`s by calling the function
`IntegerRelation::removeRedundantConstraints`. While this is only a
slight optimization in the current version, it will be necessary for
patches to come.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121989
This support has never really worked well, and is incredibly clunky to
use (it effectively creates two argument APIs), and clunky to generate (it isn't
clear how we should actually expose this from PDL frontends). Treating these
as just attribute arguments is much much cleaner in every aspect of the stack.
If we need to optimize lots of constant parameters, it would be better to
investigate internal representation optimizations (e.g. batch attribute creation),
that do not affect the user (we want a clean external API).
Differential Revision: https://reviews.llvm.org/D121569
This commit adds signature support to the language server,
and initially supports providing help for: operation operands and results,
and constraint/rewrite calls.
Differential Revision: https://reviews.llvm.org/D121545
This commit adds code completion support to the language server,
and initially supports providing completions for: Member access,
attributes/constraint/dialect/operation names, and pattern metadata.
Differential Revision: https://reviews.llvm.org/D121544
This adds support for documenting the top-level "symbols",
e.g. patterns, constraints, rewrites, etc., within a PDLL file.
Differential Revision: https://reviews.llvm.org/D121543
This adds support for providing information when hovering over
operation names, variables, patters, constraints, and rewrites.
Differential Revision: https://reviews.llvm.org/D121542
This commits adds a basic language server for PDLL to enable providing
language features in IDEs such as VSCode. This initial commit only
adds support for tracking definitions, references, and diagnostics, but
followup commits will build upon this to provide more significant behavior.
In addition to the server, this commit also updates mlir-vscode to support
the PDLL language and invoke the server.
Differential Revision: https://reviews.llvm.org/D121541
This patch slightly updates the behavior of scf.if->select to
place any hoisted select statements prior to the remaining scf.if body.
This allows better composition with other canonicalization passes, such as
scf.if nested merging.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D122027
ExpandShapeOp builder cannot infer the result type since it doesn't know
how the dimension needs to be split. Remove this builder so that it
doesn't get used accidently. Also remove one potential path using it in
generic fusion.
Differential Revision: https://reviews.llvm.org/D122019
If a stack allocation is within a nested allocation scope
don't count that as an allocation of the outer allocation scope
that would prevent inlining.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D121981
This patch extends the existing combine nested if
combination canonicalization to also handle ifs which
yield values
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D121923
Previously, the canonicalizer to create ifs from selects would only work
if the if did not have a body other than yielding. This patch upgrade the functionality
to be able to create selects from any if result whose operands are not defined
within the body.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D121943
This patch refactors the current coalesce implementation. It introduces
the `SetCoalescer`, a class in which all coalescing functionality lives.
The main advantage over the old design is the fact that the vectors of
constraints do not have to be passed around, but are implemented as
private fields of the SetCoalescer. This will become especially
important once more inequality types are introduced.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121364
This reverts commit dad80e9710.
The build is broken with some configurations (gcc-5 and gcc-8):
mlir/lib/Analysis/Presburger/PresburgerRelation.cpp:402:32: error: qualified name does not name a class before '{' token
class presburger::SetCoalescer {
This patch refactors the current coalesce implementation. It introduces
the `SetCoalescer`, a class in which all coalescing functionality lives.
The main advantage over the old design is the fact that the vectors of
constraints do not have to be passed around, but are implemented as
private fields of the SetCoalescer. This will become especially
important once more inequality types are introduced.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121364
Fold affine.load ops on global constant memrefs when indices are all
constant.
Reviewed By: ayzhuang
Differential Revision: https://reviews.llvm.org/D120612
When the sparse_tensor dialect lowers linalg.generic,
it makes inferences about how the operations should
affect the looping logic. For example, multiplication
is an intersection while addition is a union of two
sparse tensors.
The new binary and unary op separate the looping logic
from the computation by nesting the computation code
inside a block which is merged at the appropriate level
in the lowered looping code.
The binary op can have custom computation code for the
overlap, left, and right sparse overlap regions. The
unary op can have custom computation code for the
present and absent values.
Reviewed by: aartbik
Differential Revision: https://reviews.llvm.org/D121018
Fold away empty loops that iterate at least once and only return
values defined outside of the loop.
Reviewed By: bondhugula, dcaballe
Differential Revision: https://reviews.llvm.org/D121148
This allows for sharing the implementation of key components across multiple
MLIR language servers. These will be used in a followup to help implement
a PDLL language server.
Differential Revision: https://reviews.llvm.org/D121540
The current dialect registry allows for attaching delayed interfaces, that are added to attrs/dialects/ops/etc.
when the owning dialect gets loaded. This is clunky for quite a few reasons, e.g. each interface type has a
separate tracking structure, and is also quite limiting. This commit refactors this delayed mutation of
dialect constructs into a more general DialectExtension mechanism. This mechanism is essentially a registration
callback that is invoked when a set of dialects have been loaded. This allows for attaching interfaces directly
on the loaded constructs, and also allows for loading new dependent dialects. The latter of which is
extremely useful as it will now enable dependent dialects to only apply in the contexts in which they
are necessary. For example, a dialect dependency can now be conditional on if a user actually needs the
interface that relies on it.
Differential Revision: https://reviews.llvm.org/D120367
This PR exposes the region-based isTopLevelValue,
which is useful for other code that performs Affine transformations,
but is not within AffineOps.cpp
Reviewed By: bondhugula
Differential Revision: https://reviews.llvm.org/D121877
The getAffineScope function is currently internal
to AffineOps.cpp. However, as the comment on the function
itself notes, this is useful in a variety of other places
externally. This PR allows other files to use the function.
Reviewed By: bondhugula
Differential Revision: https://reviews.llvm.org/D121827
This removes any potential confusion with the `getType` accessors
which correspond to SSA results of an operation, and makes it
clear what the intent is (i.e. to represent the type of the function).
Differential Revision: https://reviews.llvm.org/D121762
This commit moves FuncOp out of the builtin dialect, and into the Func
dialect. This move has been planned in some capacity from the moment
we made FuncOp an operation (years ago). This commit handles the
functional aspects of the move, but various aspects are left untouched
to ease migration: func::FuncOp is re-exported into mlir to reduce
the actual API churn, the assembly format still accepts the unqualified
`func`. These temporary measures will remain for a little while to
simplify migration before being removed.
Differential Revision: https://reviews.llvm.org/D121266
Add basic C API for the ControlFlow dialect. Follows the format of the other dialects.
Reviewed By: mehdi_amini
Differential Revision: https://reviews.llvm.org/D121867
The current decision of when to run the verifier is running on the
assumption that nested passes can't affect the validity of the parent
operation, which isn't true. Parent operations may attach any number
of constraints on nested operations, which may not necessarily be
captured (or shouldn't be captured) at a smaller granularity.
This commit rectifies this by properly running the verifier after an
OpToOpAdaptor pass. To avoid an explosive increase in compile time,
we only run verification on the parent operation itself. To do this, a
flag to mlir::verify is added to avoid recursive verification if it isn't
desired.
Fixes#54288
Differential Revision: https://reviews.llvm.org/D121836
dense<...> expects ... to be a tensor-literal.
Define this in the grammar in BuiltinAttributes.td,
and reflect this in the reference grammar written in
AttributeParser.cpp.
Reviewed By: rriddle
Differential Revision: https://reviews.llvm.org/D121048
This removes a restriction wrt. scf.for loops during One-Shot Bufferization. Such IR was previously rejected. It is still rejected by default because the bufferized IR could be slow. But such IR can now be bufferized with `allow-return-allocs`.
Differential Revision: https://reviews.llvm.org/D121529
New buffer allocations can now be returned/yielded from blocks with `allow-return-allocs`. One-Shot Bufferize deallocates all buffers at the end of the block. If this is not possible (because the buffer escapes the block), this is now done by the existing BufferDeallocation pass.
Differential Revision: https://reviews.llvm.org/D121527
* Implement RegionBranchOpInterface: The op has a region, but it is conceptually not entered. The region just describes the semantics of the (monolithic) op.
* Linalg structured ops do not allocate memory.
Differential Revision: https://reviews.llvm.org/D121798
Such IR is rejected by default, but can be allowed with `allow-return-memref`. In preparation of future refactorings, do not deallocate such buffers.
One-Shot Analysis now gathers information about yielded tensors, so that we know during the actual bufferization whether a newly allocated buffer should be deallocated again. (Otherwise, it will leak. This will be addressed in a subsequent commit that also makes `allow-return-memref` a non-experimental flag.)
As a cleanup, `allow-return-memref` is now part of OneShotBufferizationOptions. (It was previously ignored by AlwaysCopyBufferizationState.) Moreover, AlwaysCopyBufferizationState now asserts that `create-deallocs` is deactivated to prevent surprising behavior.
Differential Revision: https://reviews.llvm.org/D121521
Defining our own function operation allows for the PDL interpreter
to be more self contained, and also removes any dependency on FuncOp;
which is moving out of the Builtin dialect.
Differential Revision: https://reviews.llvm.org/D121253
During MLIR translation to LLVMIR if an inlineable call has an UnkownLoc we get this error message:
```
inlinable function call in a function with debug info must have a !dbg location
call void @callee()
```
There is code that checks for this case and strips debug information to avoid this situation. I'm expanding this code to handle the case where an debug location points at a UnknownLoc. For example, a NamedLoc whose child location is an UnknownLoc.
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D121633
Fold linalg.fill into linalg.generic.
Remove dead arguments used in linalg.generic.
Reviewed By: ThomasRaoux
Differential Revision: https://reviews.llvm.org/D121535
When using `--convert-func-to-llvm=emit-c-wrappers` the attribute arguments of the wrapper would not be created correctly in some cases.
This patch fixes that and introduces a set of tests for (hopefully) all corner cases.
See https://github.com/llvm/llvm-project/issues/53503
Author: Sam Carroll <sam.carroll@lmns.com>
Co-Author: Laszlo Kindrat <laszlo.kindrat@lmns.com>
Reviewed By: ftynse
Differential Revision: https://reviews.llvm.org/D119895
Patch adds a new operation for the SIMD construct. The op is designed to be very similar to the existing `wsloop` operation, so that the `CanonicalLoopInfo` of `OpenMPIRBuilder` can be used.
Reviewed By: shraiysh
Differential Revision: https://reviews.llvm.org/D118065
This improves the modularity of the bufferization.
From now on, all ops that do not implement BufferizableOpInterface are considered hoisting barriers. Previously, all ops that do not implement the interface were not considered barriers and such ops had to be marked as barriers explicitly. This was unsafe because we could've hoisted across unknown ops where it was not safe to hoist.
As a side effect, this allows for cleaning up AffineBufferizableOpInterfaceImpl. This build unit no longer needed and can be deleted.
Differential Revision: https://reviews.llvm.org/D121519
Also add a TODO to switch to a custom walk instead of the GreedyPatternRewriter, which should be more efficient. (The bufferization pattern is guaranteed to apply only a single time for every op, so a simple walk should suffice.)
We currently specify a top-to-bottom walk order. This is important because other walk orders could introduce additional casts and/or buffer copies. These canonicalize away again, but it is more efficient to never generate them in the first place.
Note: A few of these canonicalizations are not yet implemented.
Differential Revision: https://reviews.llvm.org/D121518
There is currently an awkwardly complex set of rules for how a
parser/printer is generated for AttrDef/TypeDef. It can change depending on if a
mnemonic was specified, if there are parameters, if using the assemblyFormat, if
individual parser/printer code blocks were specified, etc. This commit refactors
this to make what the attribute/type wants more explicit, and to better align
with how formats are specified for operations.
Firstly, the parser/printer code blocks are removed in favor of a
`hasCustomAssemblyFormat` bit field. This aligns with the operation format
specification (and is nice to remove code blocks from ODS).
This commit also adds a requirement to explicitly set `assemblyFormat` or
`hasCustomAssemblyFormat` when the mnemonic is set and the attr/type
has no parameters. This removes the weird implicit matrix of behavior,
and also encourages the author to make a conscious choice of either C++
or declarative format instead of implicitly opting them into the C++
format (we should be pushing towards declarative when possible).
Differential Revision: https://reviews.llvm.org/D121505
OpBase.td has formed into a huge monolith of all ODS constructs. This
commits starts to rectify that by splitting out some constructs to their
own .td files.
Differential Revision: https://reviews.llvm.org/D118636
Implement the vectorLoopUnroll interface for MultiDimReduceOp and add a
pattern to do the unrolling following the same interface other vector
unroll patterns.
Differential Revision: https://reviews.llvm.org/D121263
The revision removes the linalg.fill operation and renames the OpDSL generated linalg.fill_tensor operation to replace it. After the change, all named structured operations are defined via OpDSL and there are no handwritten operations left.
A side-effect of the change is that the pretty printed form changes from:
```
%1 = linalg.fill(%cst, %0) : f32, tensor<?x?xf32> -> tensor<?x?xf32>
```
changes to
```
%1 = linalg.fill ins(%cst : f32) outs(%0 : tensor<?x?xf32>) -> tensor<?x?xf32>
```
Additionally, the builder signature now takes input and output value ranges as it is the case for all other OpDSL operations:
```
rewriter.create<linalg::FillOp>(loc, val, output)
```
changes to
```
rewriter.create<linalg::FillOp>(loc, ValueRange{val}, ValueRange{output})
```
All other changes remain minimal. In particular, the canonicalization patterns are the same and the `value()`, `output()`, and `result()` methods are now implemented by the FillOpInterface.
Depends On D120726
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D120728
Introduce an explicit `replaceOp` call to enable the tracking of the producer LinalgOp.
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D121369
This patch adds supports for union of relations (PresburgerRelation). Along
with this, support for PresburgerSet is also maintained.
This patch is part of a series of patches to add support for relations in
Presburger library.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121417
mlir-translate and related tools currently have a fixed set
of flags that are built into Translation.cpp. This works for
simple cases, but some clients want to change the default
globally (e.g. default to allowing unregistered dialects
without a command line flag), or support dialect-independent
translations without having those translations register every
conceivable dialect they could be used with (breaking
modularity).
This approach could also be applied to mlirOptMain to reduce
the significant number of flags it has accumulated.
Differential Revision: https://reviews.llvm.org/D120970
* It doesn't required by OpenCL/Intel Level Zero and can be set programmatically.
* Add GPU to spirv lowering in case when attribute is not present.
* Set higher benefit to WorkGroupSizeConversion pattern so it will always try to lower first from the attribute.
Differential Revision: https://reviews.llvm.org/D120399
In this CL, update the function name of verifier according to the
behavior. If a verifier needs to access the region then it'll be updated
to `verifyRegions`.
Reviewed By: rriddle
Differential Revision: https://reviews.llvm.org/D120373
This patch removes an old recursive implementation to lower vector.transpose to extract/insert operations
and replaces it with a iterative approach that leverages newer linearization/delinearization utilities.
The patch should be NFC except by the order in which the extract/insert ops are generated.
Reviewed By: nicolasvasilache
Differential Revision: https://reviews.llvm.org/D121321
Add operations abs, ceil, floor, and neg to the C++ API and Python API.
Add test cases.
Reviewed By: gysit
Differential Revision: https://reviews.llvm.org/D121339
This patch remove `spaceKind` from PresburgerSpace, making PresburgerSpace only
a space supporting relations.
Sets are still implemented in the same way, i.e. with a zero domain but instead
the asserts to check if the space is still set are added to users of
PresburgerSpace which treat it as a Set space.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121357
The enableObjectCache option was added in
https://reviews.llvm.org/rG06e8101034e, defaulting to false. However,
the init code added there got its logic reversed
(cache(enableObjectCache ? nullptr : new SimpleObjectCache()), which was
fixed in https://reviews.llvm.org/rGd1186fcb04 by setting the default to
true, thereby preserving the existing behavior even if it was
unintentional.
Default now the object cache to false as it was originally intended.
While at it, mention in enableObjectCache's documentation how the
cache can be dumped.
Reviewed-by: mehdi_amini
Differential Revision: https://reviews.llvm.org/D121291
This patch adds lowering from omp.atomic.update to LLVM IR. Whenever a
special LLVM IR instruction is available for the operation, `atomicrmw`
instruction is emitted, otherwise a compare-exchange loop based update
is emitted.
Depends on D119522
Reviewed By: ftynse, peixin
Differential Revision: https://reviews.llvm.org/D119657
When `addCoalescedPolyhedron` was called with `j == n - 1`,
the `polyhedrons`-vector was not properly updated (the
`IntegerPolyhedron` at position `n - 2` was "lost"). This patch adds
special handling to that case and a regression testcase.
Reviewed By: Groverkss
Differential Revision: https://reviews.llvm.org/D121356
This patch moves PresburgerSpace::removeIdRange(idStart, idLimit) to
PresburgerSpace::removeIdRange(kind, idStart, idLimit), i.e. identifiers
can only be removed at once for a single kind.
This makes users of PresburgerSpace to not assume any inside ordering of
identifier kinds.
Reviewed By: arjunp
Differential Revision: https://reviews.llvm.org/D121079
NFC. Clean up memref utils library. This library had a single function
that was completely misplaced. MemRefUtils is expected to be (also per
its comment) a library providing analysis/transforms utilities on memref
dialect ops or memref types. However, in reality it had a helper that
was depended upon by the MemRef dialect, i.e., it was a helper for the
dialect ops library and couldn't contain anything that itself depends on
the MemRef dialect. Move the single method to the memref dialect that
will now allow actual utilities depending on the memref dialect to be
placed in it.
Put findDealloc in the `memref` namespace. This is a pure move.
Differential Revision: https://reviews.llvm.org/D121273
Currently when we fold an empty loop, we assume that any loop
with iterArgs returns its iterArgs in order, which is not always
the case. It may return values defined outside of the loop or
return its iterArgs out of order. This patch adds support to
those cases.
Reviewed By: dcaballe
Differential Revision: https://reviews.llvm.org/D120776
This revision adds support for the linalg.index to the sparse compiler
pipeline. In essence, this adds the ability to refer to indices in
the tensor index expression, as illustrated below:
Y[i, j, k, l, m] = T[i, j, k, l, m] * i * j
Reviewed By: bixia
Differential Revision: https://reviews.llvm.org/D121251
BuiltinOps.h
These includes are going to be removed from BuiltinOps.h in a followup
when FuncOp is moved out of the Builtin dialect. This commit
pre-emptively adds those includes to simplify the patch moving FuncOp.
This feels like a layering violation, but it fixes the build.
Fixes#54242
tools/mlir/lib/Dialect/GPU/CMakeFiles/obj.MLIRGPUTransforms.dir/Transforms/SerializeToHsaco.cpp.o:SerializeToHsaco.cpp:function (anonymous namespace)::SerializeToHsacoPass::optimizeLlvm(llvm::Module&, llvm::TargetMachine&):
error: undefined reference to 'mlir::makeOptimizingTransformer(unsigned int, unsigned int, llvm::TargetMachine*)'
It is currently a module pass, but shouldn't be. All of the patterns
are local conversions, and don't require anything about
functions/modules.
Differential Revision: https://reviews.llvm.org/D121192
River Riddle