summaryrefslogtreecommitdiffstats
path: root/mlir/test/Dialect/LLVMIR
Commit message (Collapse)AuthorAgeFilesLines
* Automated rollback of commit f68ac464d818629e0fe10c23b44ac782d64a12d2Christian Sigg2019-12-122-11/+11
| | | | PiperOrigin-RevId: 285162061
* Switch from shfl.bfly to shfl.down.Christian Sigg2019-12-122-11/+11
| | | | | | | 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
* LLVM::AddressOfOp: properly take into account the address spaceAlex Zinenko2019-12-061-0/+16
| | | | | | | | | | The AddressOf operation in the LLVM dialect return a pointer to a global variable. The latter may be in a non-default address space as indicated by the "addr_space" attribute. Check that the address space of the pointer returned by AddressOfOp matches that of the referenced GlobalOp. Update the AddressOfOp builder to respect this constraint. PiperOrigin-RevId: 284138860
* Add linkage support to LLVMFuncOpAlex Zinenko2019-12-031-0/+35
| | | | | | | | | A recent commit introduced the Linkage attribute to the LLVM dialect and used it in the Global Op. Also use it in LLVMFuncOp. As per LLVM Language Reference, if the linkage attribute is omitted, the function is assumed to have external linkage. PiperOrigin-RevId: 283493299
* Introduce Linkage attribute to the LLVM dialectAlex Zinenko2019-12-021-24/+46
| | | | | | | | | | | LLVM IR supports linkage on global objects such as global variables and functions. Introduce the Linkage attribute into the LLVM dialect, backed by an integer storage. Use this attribute on LLVM::GlobalOp and make it mandatory. Implement parsing/printing of the attribute and conversion to LLVM IR. See tensorflow/mlir#277. PiperOrigin-RevId: 283309328
* Allow LLVM::ExtractElementOp to have non-i32 indices.MLIR Team2019-11-252-4/+4
| | | | | | Also change the text format a bit, so that indices are braced by squares. PiperOrigin-RevId: 282437095
* Don't force newline before function attributesAlex Zinenko2019-11-211-4/+4
| | | | | | | | | | | Due to legacy reasons, a newline character followed by two spaces was always inserted before the attributes of the function Op in pretty form. This breaks formatting when functions are nested in some other operations. Don't print the newline and just put the attributes on the same line, which is also more consistent with module Op. Line breaking aware of indentation can be introduced separately into the parser if deemed useful. PiperOrigin-RevId: 281721793
* Add support for the LLVM FNeg instructionEric Schweitz2019-11-061-0/+3
| | | | | | | Closes tensorflow/mlir#216 COPYBARA_INTEGRATE_REVIEW=https://github.com/tensorflow/mlir/pull/216 from schweitzpgi:llvmir-fneg-op f9b5f185845d671b745ab6fc213d5d9aff044b34 PiperOrigin-RevId: 278795325
* [llvm] Allow GlobalOp to take a region for complex initializersJames Molloy2019-11-051-0/+33
| | | | | | | | | | | | | | | This allows GlobalOp to either take a value attribute (for simple constants) or a region that can contain IR instructions (that must be constant-foldable) to create a ConstantExpr initializer. Example: // A complex initializer is constructed with an initializer region. llvm.mlir.global constant @int_gep() : !llvm<"i32*"> { %0 = llvm.mlir.addressof @g2 : !llvm<"i32*"> %1 = llvm.mlir.constant(2 : i32) : !llvm.i32 %2 = llvm.getelementptr %0[%1] : (!llvm<"i32*">, !llvm.i32) -> !llvm<"i32*"> llvm.return %2 : !llvm<"i32*"> } PiperOrigin-RevId: 278717836
* [NVVM] Add mma.sync operation.MLIR Team2019-11-042-0/+105
| | | | PiperOrigin-RevId: 278440547
* [mlir][llvm] Add missing cast opsJames Molloy2019-11-011-1/+11
| | | | | | Also adds a builder method for fcmp, identical to that for icmp. PiperOrigin-RevId: 277923158
* Convert the Canonicalize and CSE passes to generic Operation Passes.River Riddle2019-10-241-1/+1
| | | | | | This allows for them to be used on other non-function, or even other function-like, operations. The algorithms are already generic, so this is simply changing the derived pass type. The majority of this change is just ensuring that the nesting of these passes remains the same, as the pass manager won't auto-nest them anymore. PiperOrigin-RevId: 276573038
* Add a Symbol trait to simplify defining operations that represent symbols.River Riddle2019-10-211-1/+1
| | | | | | This trait provides accessors for the name, symbol use list methods, verification, with more to be added. PiperOrigin-RevId: 275864554
* Fix minor spelling tweaks (NFC)Kazuaki Ishizaki2019-10-202-3/+3
| | | | | | Closes tensorflow/mlir#175 PiperOrigin-RevId: 275726876
* Get active source lane predicate from shuffle instruction.Christian Sigg2019-10-192-0/+33
| | | | | | | | 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
* Allow '_' when pretty printing dialect symbolsLei Zhang2019-10-171-2/+2
| | | | | | | | | '_' is used frequently enough as the separator of words in symbols. We should allow it in dialect symbols when considering pretty printing. Also updated LangRef.md regarding pretty form. PiperOrigin-RevId: 275312494
* Add LLVM IR dialect hooks for FP128 and X86_FP80 typesEric Schweitz2019-10-111-0/+7
| | | | | | | Closes tensorflow/mlir#184 COPYBARA_INTEGRATE_REVIEW=https://github.com/tensorflow/mlir/pull/184 from schweitzpgi:more-float-types ca27d00510a86ffc9c79c65fb3a0193b5ea097a0 PiperOrigin-RevId: 274288813
* LLVM Dialect: introduce llvm.mlir.null operationAlex Zinenko2019-10-112-0/+17
| | | | | | | | Similarly to `llvm.mlir.undef`, this auxiliary operation creates an SSA value that corresponds to `null` in LLVM IR. This operation is necessary to model sizeof(<...>) behavior when allocating memory. PiperOrigin-RevId: 274158760
* [ROCm] Adding ROCDL Dialect.Deven Desai2019-09-271-0/+30
| | | | | | | | | | | | | This commit introduces the ROCDL Dialect (i.e. the ROCDL ops + the code to lower those ROCDL ops to LLWM intrinsics/functions). Think of ROCDL Dialect as analogous to the NVVM Dialect, but for AMD GPUs. This patch contains just the essentials needed to get a simple example up and running. We expect to make further additions to the ROCDL Dialect. This is the first of 3 commits, the follow-up will be: * add a pass that lowers GPU Dialect to ROCDL Dialect * add a "mlir-rocm-runner" utility Closes tensorflow/mlir#146 COPYBARA_INTEGRATE_REVIEW=https://github.com/tensorflow/mlir/pull/146 from deven-amd:deven-rocdl-dialect e78e8005c75a78912631116c78dc844fcc4b0de9 PiperOrigin-RevId: 271511259
* Make GlobalOp's value attribute optional.Christian Sigg2019-09-211-5/+3
| | | | | | Make GlobalOp's value attribute an OptionalAttr. Change code that uses the value to handle 'nullopt'. Translate an unitialized value attribute to llvm::UndefValue. PiperOrigin-RevId: 270423646
* Add address space attribute to LLVMIR's GlobalOp.MLIR Team2019-09-191-0/+10
| | | | PiperOrigin-RevId: 270012505
* Move LLVMIR dialect tests from test/LLVMIR to test/Dialect and test/ConversionAlex Zinenko2019-09-046-0/+810
This follows up on the recent restructuring that moved the dialects under lib/Dialect and inter-dialect conversions to lib/Conversion. Originally, the tests for both the LLVMIR dialect itself and the conversion from Standard to LLVMIR dialect lived under test/LLVMIR. This no longer reflects the code structure. Move the tests to either test/Dialect/LLVMIR or test/Conversion/StandardToLLVM depending on the features they exercise. PiperOrigin-RevId: 267159219
OpenPOWER on IntegriCloud