31 Commits

Author SHA1 Message Date
Mehdi Amini f164534ca8 Add a dialect_registration callback for "translations" registered with mlir-translate
This will allow out-of-tree translation to register the dialects they expect
to see in their input, on the model of getDependentDialects() for passes.

Differential Revision: https://reviews.llvm.org/D86409
2020-08-23 01:00:39 +00:00
Alex Zinenko db1c197bf8 [mlir] take LLVMContext in MLIR-to-LLVM-IR translation
Due to the original type system implementation, LLVMDialect in MLIR contains an
LLVMContext in which the relevant objects (types, metadata) are created. When
an MLIR module using the LLVM dialect (and related intrinsic-based dialects
NVVM, ROCDL, AVX512) is converted to LLVM IR, it could only live in the
LLVMContext owned by the dialect. The type system no longer relies on the
LLVMContext, so this limitation can be removed. Instead, translation functions
now take a reference to an LLVMContext in which the LLVM IR module should be
constructed. The caller of the translation functions is responsible for
ensuring the same LLVMContext is not used concurrently as the translation no
longer uses a dialect-wide context lock.

As an additional bonus, this change removes the need to recreate the LLVM IR
module in a different LLVMContext through printing and parsing back, decreasing
the compilation overhead in JIT and GPU-kernel-to-blob passes.

Reviewed By: rriddle, mehdi_amini

Differential Revision: https://reviews.llvm.org/D85443
2020-08-07 14:22:30 +02:00
Jonathan Roelofs 223154d267 [mlir] Remove need for static global ctors from mlir-translate
Summary: https://bugs.llvm.org/show_bug.cgi?id=45436

Reviewers: mehdi_amini, mravishankar, antiagainst, rriddle, stephenneuendorffer

Reviewed By: mehdi_amini, rriddle, stephenneuendorffer

Subscribers: frgossen, stephenneuendorffer, jholewinski, mgorny, rriddle, jpienaar, burmako, shauheen, antiagainst, nicolasvasilache, arpith-jacob, mgester, lucyrfox, aartbik, liufengdb, Joonsoo, bader, grosul1, llvm-commits

Tags: #llvm

Differential Revision: https://reviews.llvm.org/D77515
2020-04-08 16:52:33 -06:00
River Riddle 93dc8bd267 [mlir] Add explicit friendship to LLVM::ModuleTranslation to derived
classes.

This allows for the `LLVM::ModuleTranslation::translateModule` to properly access the constructors of the derived classes.
2020-02-05 18:34:44 -08:00
River Riddle c33d6970e0 [mlir] Add support for basic location translation to LLVM.
Summary:
This revision adds basic support for emitting line table information when exporting to LLVMIR. We don't yet have a story for supporting all of the LLVM debug metadata, so this revision stubs some features(like subprograms) to enable emitting line tables.

Differential Revision: https://reviews.llvm.org/D73934
2020-02-05 17:41:51 -08:00
Mehdi Amini 308571074c Mass update the MLIR license header to mention "Part of the LLVM project"
This is an artifact from merging MLIR into LLVM, the file headers are
now aligned with the rest of the project.
2020-01-26 03:58:30 +00:00
Mehdi Amini 56222a0694 Adjust License.txt file to use the LLVM license
PiperOrigin-RevId: 286906740
2019-12-23 15:33:37 -08:00
River Riddle 2666b97314 NFC: Cleanup non-conforming usages of namespaces.
* Fixes use of anonymous namespace for static methods.
* Uses explicit qualifiers(mlir::) instead of wrapping the definition with the namespace.

PiperOrigin-RevId: 286222654
2019-12-18 10:46:48 -08:00
River Riddle 4562e389a4 NFC: Remove unnecessary 'llvm::' prefix from uses of llvm symbols declared in mlir namespace.
Aside from being cleaner, this also makes the codebase more consistent.

PiperOrigin-RevId: 286206974
2019-12-18 09:29:20 -08:00
MLIR Team 6e581e29a4 Integrate from upstream at revision e4fce659a7.
PiperOrigin-RevId: 285982330
2019-12-17 08:13:14 -08:00
Tres Popp 44fc7d72b3 Remove LLVM dependency on mlir::Module and instead check Traits.
PiperOrigin-RevId: 285724678
2019-12-16 01:45:44 -08:00
Christian Sigg 9b85582682 Automated rollback of commit f68ac464d8
PiperOrigin-RevId: 285162061
2019-12-12 03:48:38 -08:00
Christian Sigg f68ac464d8 Switch from shfl.bfly to shfl.down.
Both work for the current use case, but the latter allows implementing
prefix sums and is a little easier to understand for partial warps.

PiperOrigin-RevId: 285145287
2019-12-12 01:28:01 -08:00
Christian Sigg c3e56cd12c Get active source lane predicate from shuffle instruction.
nvvm.shfl.sync.bfly optionally returns a predicate whether source lane was active. Support for this was added to clang in https://reviews.llvm.org/D68892.

Add an optional 'pred' unit attribute to the instruction to return this predicate. Specify this attribute in the partial warp reduction so we don't need to manually compute the predicate.

PiperOrigin-RevId: 275616564
2019-10-19 01:53:25 -07:00
Alex Zinenko 5e7959a353 Use llvm.func to define functions with wrapped LLVM IR function type
This function-like operation allows one to define functions that have wrapped
LLVM IR function type, in particular variadic functions. The operation was
added in parallel to the existing lowering flow, this commit only switches the
flow to use it.

Using a custom function type makes the LLVM IR dialect type system more
consistent and avoids complex conversion rules for functions that previously
had to use the built-in function type instead of a wrapped LLVM IR dialect type
and perform conversions during the analysis.

PiperOrigin-RevId: 273910855
2019-10-10 01:34:06 -07:00
Lei Zhang b00a522b80 Change MLIR translation functions signature
This CL changes translation functions to take MemoryBuffer
as input and raw_ostream as output. It is generally better to
avoid handling files directly in a library (unless the library
is specifically for file manipulation) and we can unify all
file handling to the mlir-translate binary itself.

PiperOrigin-RevId: 269625911
2019-09-17 12:16:45 -07:00
MLIR Team 696fcb7520 Add 3 additional intrinsic ops to NVVM dialect, in preparation to implement block-wide reduce.
PiperOrigin-RevId: 265720077
2019-08-27 10:56:18 -07:00
River Riddle ba0fa92524 NFC: Move LLVMIR, SDBM, and StandardOps to the Dialect/ directory.
PiperOrigin-RevId: 264193915
2019-08-19 11:01:25 -07:00
Alex Zinenko baa1ec22f7 Translation to LLVM IR: use LogicalResult instead of bool
The translation code predates the introduction of LogicalResult and was relying
on the obsolete LLVM convention of returning false on success.  Change it to
use MLIR's LogicalResult abstraction instead. NFC.

PiperOrigin-RevId: 262589432
2019-08-09 10:45:44 -07:00
Alex Zinenko 60965b4612 Move GPU dialect to {lib,include/mlir}/Dialect
Per tacit agreement, individual dialects should now live in lib/Dialect/Name
with headers in include/mlir/Dialect/Name and tests in test/Dialect/Name.

PiperOrigin-RevId: 259896851
2019-07-25 00:41:17 -07:00
River Riddle fec20e590f NFC: Rename Module to ModuleOp.
Module is a legacy name that only exists as a typedef of ModuleOp.

PiperOrigin-RevId: 257427248
2019-07-10 10:11:21 -07:00
River Riddle 8c44367891 NFC: Rename Function to FuncOp.
PiperOrigin-RevId: 257293379
2019-07-10 10:10:53 -07:00
River Riddle 626b8b6a5d NFC: Remove Module::getFunctions in favor of a general getOps<T>.
Modules can now contain more than just Functions, this just updates the iteration API to reflect that. The 'begin'/'end' methods have also been updated to iterate over opaque Operations.

PiperOrigin-RevId: 257099084
2019-07-08 18:28:17 -07:00
Alex Zinenko 6f7f2bced8 Make TranslateFromMLIRFunction type return LogicalResult instead of bool
This interface was created before MLIR introduced LogicalResult.

PiperOrigin-RevId: 256554557
2019-07-04 08:19:14 -07:00
River Riddle e7d594bb1c Replace the implementation of Function and Module with FuncOp and ModuleOp.
This is an important step in allowing for the top-level of the IR to be extensible. FuncOp and ModuleOp contain all of the necessary functionality, while using the existing operation infrastructure. As an interim step, many of the usages of Function and Module, including the name, will remain the same. In the future, many of these will be relaxed to allow for many different types of top-level operations to co-exist.

PiperOrigin-RevId: 256427100
2019-07-03 14:37:18 -07:00