This revision adds a pre-bufferization transform that can reduce the number of allocation. It is similar to `bufferization.eliminate_empty_tensors`, but specific to LinalgOp.
The transform looks for `tensor.empty` ops where the SSA use-def chain ends in an "ins" operand of a `LinalgOp`. If the same `LinalgOp` has an unused "outs" operand (and some other conditions are met), this "outs" operand can be used instead of the `tensor.empty` and the "ins" operand can be turned into an "outs" operand.
Differential Revision: https://reviews.llvm.org/D153952
This is based on ideas from @nafi to:
- use a branchless version of 'cmp' for 'uint32_t',
- completely resolve the lexicographic comparison through vector
operations when wide types are available. We also get rid of byte
reloads and serializing '__builtin_ctzll'.
I did not include the suggestion to replace comparisons of 'uint16_t'
with two 'uint8_t' as it did not seem to help the codegen. This can
be revisited in sub-sequent patches.
The code been rewritten to reduce nested function calls, making the
job of the inliner easier and preventing harmful code duplication.
Reviewed By: nafi3000
Differential Revision: https://reviews.llvm.org/D148717
We are in the progress of migrating to a much improved surface syntax for the Sparse Tensor Encoding Attribute (STEA).
You can see a preview of this in the StableHLO RFC at
https://github.com/openxla/stablehlo/blob/main/rfcs/20230210-sparsity.md
//**This design is courtesy Wren Romano.**//
This initial revision
(1) Introduces the first version of a new parser written by Wren Romano
(2) Introduces a simple "migration plan" using NEW_SYNTAX on the STEA, which will allow us to test the new parser with new examples, as well as migrate existing examples over without the need to rewrite them all
This first "drop" merely provides the entry points to parse the new syntax. The parser is still under active development. For example, we need to address the "lookahead" issue when parsing the lvl spec (viz. do we see l0 = d0 or a direct d0). Another larger task is to actually implement "affine" parsing (since the MLIR affine parser is not accessible in other parts of the tree).
EXAMPLE:
Currently, CSR looks like
#CSR = #sparse_tensor.encoding<{
lvlTypes = ["dense","compressed"],
dimToLvl = affine_map<(i,j) -> (i,j)>
}>
but you can "force" the new parser with
#CSR = #sparse_tensor.encoding<{
NEW_SYNTAX =
(d0, d1) -> (l0 = d0 : dense, l1 = d1 : compressed)
}>
Reviewed By: Peiming
Differential Revision: https://reviews.llvm.org/D153997
Switch the parse of command line options fromllvm::cl to OptTable.
The motivation for this change is to continue adding llvm based tools
to the llvm driver multicall.
Differential Revision: https://reviews.llvm.org/D153665
Mapping to NVGPU operations such as mma.sync with mixed precision and ldmatrix with transposes and
various data types involves complex matchings from low-level IR.
This is akin to raising complex patterns after unnecessarily having lost structural information.
To avoid such unnecessary complexity, introduce a direct mapping step from a matmul on memrefs
to distributed NVGPU vector abstractions.
In this context, mapping to specific mma.sync operations is trivial and consists in simply
translating the documentation into indexing expressions.
Correctness is demonstrated with an end-to-end integration test.
Differential Revision: https://reviews.llvm.org/D153420
This change also prepares for 9119325 to land again.
Adds `mlir_c_runner_utils_hdrs` and `mlir_runner_utils_hdrs` targets which do not depend on `//llvm::Support`.
These can be used by other 'runner.so' targets if they are loaded along with the 'runner_utils.so' without calling `__mlir_execution_engine_init()` twice.
Once integrated in our codebase the patch triggered a bunch of failing
tests. We do not yet understand where the bug is but we revert it to
move forward with integration.
This reverts commit 5e32765c15.
The Bazel build was relying, for the two files enumerated in this diff, on the legacy implicit-export semantics described here:
https://bazel.build/reference/be/functions#exports_files
This documentation page encourages migrating away from this legacy behavior, and indeed we have a user who reported a Bazel build error and it appears that they were already using the new, stricter behavior:
https://github.com/openxla/iree/pull/13982
and while examining fixes on our side and trying to get a clean Bazel build, I ran into this similar issue in the LLVM overlay.
It would arguably be cleaner (in the sense of more structured) to rely on `filegroup` to export this, but I am insufficiently familiar with the Clang build (the dependent targets seem to be below Clang) to do this myself. The present `exports_files` solution has the merit of being localized in these few lines here.
Differential Revision: https://reviews.llvm.org/D152491
Fixing an issue with LLVM libc's fenv.h defined rounding mode macros
differently from system libc, making get_round() return different values from
fegetround(). Also letting math tests to skip rounding modes that cannot be
set. This should allow math tests to be run on platforms in which fenv.h is not
implemented yet.
This allows us to re-enable hermatic floating point tests in
https://reviews.llvm.org/D151123 and reverting https://reviews.llvm.org/D152742.
Reviewed By: jhuber6
Differential Revision: https://reviews.llvm.org/D152873
This patch mimics the behavior of Google Test and allow users to log custom messages after all flavors of ASSERT_ / EXPECT_.
Reviewed By: sivachandra, lntue
Differential Revision: https://reviews.llvm.org/D152630
This patch mimics the behavior of Google Test and allow users to log custom messages after all flavors of ASSERT_ / EXPECT_.
Reviewed By: sivachandra, lntue
Differential Revision: https://reviews.llvm.org/D152630
Add Int<> and Int128 types to replace the usage of __int128_t in math
functions. Clean up to make sure that (U)Int128 and __(u)int128_t are
interchangeable in the code base.
Reviewed By: sivachandra, mikhail.ramalho
Differential Revision: https://reviews.llvm.org/D152459
The antique version of bash (3.2.57, from 2007) which is available on
macOS cannot deal with quoted slashes in a `${x/...}`
substitution. Since only prefix-removal is required here, switch to a
`${x#...}` substitution instead.
(E.g. `src="foo/bar/baz.h"; echo ${src/"foo/bar"}` echos `bar/bar/baz.h`
instead of `/baz.h` on old bash versions).
Originally broken by 459420c33a.
Fixes#63222
This patch adds the reentrent qsort entrypoint, qsort_r. This is done by
extending the qsort functionality and moving it to a shared utility
header. For this reason the qsort_r tests focus mostly on the places
where it differs from qsort, since they share the same sorting code.
Reviewed By: sivachandra, lntue
Differential Revision: https://reviews.llvm.org/D152467
This is based on ideas from @nafi to:
- use a branchless version of 'cmp' for 'uint32_t',
- completely resolve the lexicographic comparison through vector
operations when wide types are available. We also get rid of byte
reloads and serializing '__builtin_ctzll'.
I did not include the suggestion to replace comparisons of 'uint16_t'
with two 'uint8_t' as it did not seem to help the codegen. This can
be revisited in sub-sequent patches.
The code been rewritten to reduce nested function calls, making the
job of the inliner easier and preventing harmful code duplication.
Reviewed By: nafi3000
Differential Revision: https://reviews.llvm.org/D148717
Many math functions need to check for floating point rounding modes to
return correct values. Currently most of them use the internal implementation
of `fegetround`, which is platform-dependent and blocking math functions to be
enabled on platforms with unimplemented `fegetround`. In this change, we add
platform independent rounding mode checks and switching math functions to use
them instead. https://github.com/llvm/llvm-project/issues/63016
Reviewed By: sivachandra
Differential Revision: https://reviews.llvm.org/D152280
This is based on ideas from @nafi to:
- use a branchless version of 'cmp' for 'uint32_t',
- completely resolve the lexicographic comparison through vector
operations when wide types are available. We also get rid of byte
reloads and serializing '__builtin_ctzll'.
I did not include the suggestion to replace comparisons of 'uint16_t'
with two 'uint8_t' as it did not seem to help the codegen. This can
be revisited in sub-sequent patches.
The code been rewritten to reduce nested function calls, making the
job of the inliner easier and preventing harmful code duplication.
Reviewed By: nafi3000
Differential Revision: https://reviews.llvm.org/D148717
str method of FPBits class is only used for pretty printing its objects
in tests. It brings cpp::string dependency to FPBits class, which is not ideal
for embedded use case. We move str method to a free function in test utils and
remove this dependency of FPBits class.
Reviewed By: sivachandra
Differential Revision: https://reviews.llvm.org/D152607
The Mega Table that printf uses for long doubles with some flags is too
large for the linters, and so has been split out from the main patch.
The main patch: https://reviews.llvm.org/D150399
Reviewed By: sivachandra
Differential Revision: https://reviews.llvm.org/D152470
This patch adds three options for printf decimal long doubles, and these
can also apply to doubles.
1. Use a giant table which is fast and accurate, but takes up ~5MB).
2. Use dyadic floats for approximations, which only gives ~50 digits of
accuracy but is very fast.
3. Use large integers for approximations, which is accurate but very
slow.
Reviewed By: sivachandra, lntue
Differential Revision: https://reviews.llvm.org/D150399
