[mlir] Set the namespace of the BuiltinDialect to 'builtin'

[rriddle]

Summary

Historically the builtin dialect has had an empty namespace. This has unfortunately created a very awkward situation, where many utilities either have to special case the empty namespace, or just don't work at all right now. This revision adds a namespace to the builtin dialect, and starts to cleanup some of the utilities to no longer handle empty namespaces. For now, the assembly form of builtin operations does not require the builtin. prefix. (This should likely be re-evaluated though)

[rriddle]

Fix python test

[ftynse]

mlir/docs/Diagnostics.md

159

Maybe irrelevant: is "module_terminator" still a thing? If so, should it also be builtin?

[mehdi_amini]

mlir/docs/Diagnostics.md

159

It isn't, we should update the doc with another example here.

flang/lib/Optimizer/Dialect/FIRType.cpp

 //===-- FIRType.cpp -------------------------------------------------------===//
 //
 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
 // See https://llvm.org/LICENSE.txt for license information.
 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
 //
 //===----------------------------------------------------------------------===//
 //
 // Coding style: https://mlir.llvm.org/getting_started/DeveloperGuide/
 //
 //===----------------------------------------------------------------------===//
 
 #include "flang/Optimizer/Dialect/FIRType.h"
 #include "flang/Optimizer/Dialect/FIRDialect.h"
 #include "mlir/IR/Builders.h"
+#include "mlir/IR/BuiltinDialect.h"
 #include "mlir/IR/Diagnostics.h"
 #include "mlir/IR/DialectImplementation.h"
 #include "llvm/ADT/SmallPtrSet.h"
 #include "llvm/ADT/StringSet.h"
 #include "llvm/ADT/TypeSwitch.h"
 #include "llvm/Support/ErrorHandling.h"
 
 #define GET_TYPEDEF_CLASSES
bool inbounds(A v, B lb, B ub) {
   return v >= lb && v < ub;
 }
 
 bool isa_fir_type(mlir::Type t) {
   return llvm::isa<FIROpsDialect>(t.getDialect());
 }
 
 bool isa_std_type(mlir::Type t) {
-  return t.getDialect().getNamespace().empty();
+  return llvm::isa<mlir::BuiltinDialect>(t.getDialect());
 }
 
 bool isa_fir_or_std_type(mlir::Type t) {
   if (auto funcType = t.dyn_cast<mlir::FunctionType>())
     return llvm::all_of(funcType.getInputs(), isa_fir_or_std_type) &&
            llvm::all_of(funcType.getResults(), isa_fir_or_std_type);
   return isa_fir_type(t) || isa_std_type(t);
 }

flang/test/Fir/cg-ops.fir

-// RUN: fir-opt --pass-pipeline="func(cg-rewrite),fir.global(cg-rewrite),cse" %s | FileCheck %s
+// RUN: fir-opt --pass-pipeline="builtin.func(cg-rewrite),fir.global(cg-rewrite),cse" %s | FileCheck %s
 
 // CHECK-LABEL: func @codegen(
 // CHECK-SAME: %[[arg:.*]]: !fir
 func @codegen(%addr : !fir.ref<!fir.array<?xi32>>) {
   // CHECK: %[[zero:.*]] = constant 0 : index
   %0 = constant 0 : index
   %1 = fir.shape_shift %0, %0 : (index, index) -> !fir.shapeshift<1>
   %2 = fir.slice %0, %0, %0 : (index, index, index) -> !fir.slice<1>

mlir/docs/Diagnostics.md

 
 When a diagnostic is emitted on an operation, via `Operation::emitError/...`,
 the textual form of that operation is printed and attached as a note to the
 diagnostic. This option is useful for understanding the current form of an
 operation that may be invalid, especially when debugging verifier failures. An
 example output is shown below:
 
 ```shell
-test.mlir:3:3: error: 'module_terminator' op expects parent op 'module'
+test.mlir:3:3: error: 'module_terminator' op expects parent op 'builtin.module'
   "module_terminator"() : () -> ()
   ^
 test.mlir:3:3: note: see current operation: "module_terminator"() : () -> ()
   "module_terminator"() : () -> ()
   ^
 ```
 
 ### Print StackTrace On Diagnostic
 
 Command Line Flag: `-mlir-print-stacktrace-on-diagnostic`
 
 When a diagnostic is emitted, attach the current stack trace as a note to the
 diagnostic. This option is useful for understanding which part of the compiler
 generated certain diagnostics. An example output is shown below:
 
 ```shell
-test.mlir:3:3: error: 'module_terminator' op expects parent op 'module'
+test.mlir:3:3: error: 'module_terminator' op expects parent op 'builtin.module'
   "module_terminator"() : () -> ()
   ^
 test.mlir:3:3: note: diagnostic emitted with trace:
  #0 0x000055dd40543805 llvm::sys::PrintStackTrace(llvm::raw_ostream&) llvm/lib/Support/Unix/Signals.inc:553:11
  #1 0x000055dd3f8ac162 emitDiag(mlir::Location, mlir::DiagnosticSeverity, llvm::Twine const&) /lib/IR/Diagnostics.cpp:292:7
  #2 0x000055dd3f8abe8e mlir::emitError(mlir::Location, llvm::Twine const&) /lib/IR/Diagnostics.cpp:304:10
  #3 0x000055dd3f998e87 mlir::Operation::emitError(llvm::Twine const&) /lib/IR/Operation.cpp:324:29
  #4 0x000055dd3f99d21c mlir::Operation::emitOpError(llvm::Twine const&) /lib/IR/Operation.cpp:652:10

mlir/docs/PassManagement.md

 Below is an example of constructing a pipeline that operates on the above
 structure:
 
 ```c++
 // Create a top-level `PassManager` class. If an operation type is not
 // explicitly specific, the default is the builtin `module` operation.
 PassManager pm(ctx);
 // Note: We could also create the above `PassManager` this way.
-PassManager pm(ctx, /*operationName=*/"module");
+PassManager pm(ctx, /*operationName=*/"builtin.module");
 
 // Add a pass on the top-level module operation.
 pm.addPass(std::make_unique<MyModulePass>());
 
 // Nest a pass manager that operates on `spirv.module` operations nested
 // directly under the top-level module.
 OpPassManager &nestedModulePM = pm.nest<spirv::ModuleOp>();
 nestedModulePM.addPass(std::make_unique<MySPIRVModulePass>());

 Pass::runPipeline(OpPassManager &, Operation *)` method. This method returns
 whether the dynamic pipeline succeeded or failed, similarly to the result of the
 top-level `PassManager::run` method. A simple example is shown below:
 
 ```c++
 void MyModulePass::runOnOperation() {
   ModuleOp module = getOperation();
   if (hasSomeSpecificProperty(module)) {
-    OpPassManager dynamicPM("module");
+    OpPassManager dynamicPM("builtin.module");
     ...; // Build the dynamic pipeline.
     if (failed(runPipeline(dynamicPM, module)))
       return signalPassFailure();
   }
 }
 ```
 
 Note: though above the dynamic pipeline was constructed within the

 ```
 
 The collected statistics may be aggregated in two types of views:
 
 A pipeline view that models the structure of the pass manager, this is the
 default view:
 
 ```shell
-$ mlir-opt -pass-pipeline='func(my-pass,my-pass)' foo.mlir -pass-statistics
+$ mlir-opt -pass-pipeline='builtin.func(my-pass,my-pass)' foo.mlir -pass-statistics
 
 ===-------------------------------------------------------------------------===
                          ... Pass statistics report ...
 ===-------------------------------------------------------------------------===
-'func' Pipeline
+'builtin.func' Pipeline
   MyPass
     (S) 15 exampleStat - An example statistic
   VerifierPass
   MyPass
     (S)  6 exampleStat - An example statistic
   VerifierPass
 VerifierPass
 ```
 
 A list view that aggregates the statistics of all instances of a specific pass
 together:
 
 ```shell
-$ mlir-opt -pass-pipeline='func(my-pass, my-pass)' foo.mlir -pass-statistics -pass-statistics-display=list
+$ mlir-opt -pass-pipeline='builtin.func(my-pass, my-pass)' foo.mlir -pass-statistics -pass-statistics-display=list
 
 ===-------------------------------------------------------------------------===
                          ... Pass statistics report ...
 ===-------------------------------------------------------------------------===
 MyPass
   (S) 21 exampleStat - An example statistic
 ```
 

 ```ebnf
 pipeline          ::= op-name `(` pipeline-element (`,` pipeline-element)* `)`
 pipeline-element  ::= pipeline | (pass-name | pass-pipeline-name) options?
 options           ::= '{' (key ('=' value)?)+ '}'
 ```
 
 *   `op-name`
     *   This corresponds to the mnemonic name of an operation to run passes on,
-        e.g. `func` or `module`.
+        e.g. `builtin.func` or `builtin.module`.
 *   `pass-name` | `pass-pipeline-name`
     *   This corresponds to the argument of a registered pass or pass pipeline,
         e.g. `cse` or `canonicalize`.
 *   `options`
     *   Options are specific key value pairs representing options defined by a
         pass or pass pipeline, as described in the
         ["Instance Specific Pass Options"](#instance-specific-pass-options)
         section. See this section for an example usage in a textual pipeline.
 
 For example, the following pipeline:
 
 ```shell
 $ mlir-opt foo.mlir -cse -canonicalize -convert-std-to-llvm='use-bare-ptr-memref-call-conv=1'
 ```
 
 Can also be specified as (via the `-pass-pipeline` flag):
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse,canonicalize),convert-std-to-llvm{use-bare-ptr-memref-call-conv=1}'
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse,canonicalize),convert-std-to-llvm{use-bare-ptr-memref-call-conv=1}'
 ```
 
 In order to support round-tripping a pass to the textual representation using
 `OpPassManager::printAsTextualPipeline(raw_ostream&)`, override `StringRef
 Pass::getArgument()` to specify the argument used when registering a pass.
 
 ## Declarative Pass Specification
 

 
 In this mode, the results are displayed in a list sorted by total time with each
 pass/analysis instance aggregated into one unique result. This view is useful
 for getting an overview of what analyses/passes are taking the most time in a
 pipeline. This display mode is available in mlir-opt via
 `-mlir-timing-display=list`.
 
 ```shell
-$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing -mlir-timing-display=list
+$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='builtin.func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing -mlir-timing-display=list
 
 ===-------------------------------------------------------------------------===
                       ... Pass execution timing report ...
 ===-------------------------------------------------------------------------===
   Total Execution Time: 0.0203 seconds
 
    ---Wall Time---  --- Name ---
    0.0047 ( 55.9%)  Canonicalizer
    0.0019 ( 22.2%)  VerifierPass
    0.0016 ( 18.5%)  LLVMLoweringPass
    0.0003 (  3.4%)  CSE
    0.0002 (  1.9%)  (A) DominanceInfo
    0.0084 (100.0%)  Total
 ```
 
 ##### Tree Display Mode
 
 In this mode, the results are displayed in a nested pipeline view that mirrors
 the internal pass pipeline that is being executed in the pass manager. This view
 is useful for understanding specifically which parts of the pipeline are taking
 the most time, and can also be used to identify when analyses are being
 invalidated and recomputed. This is the default display mode.
 
 ```shell
-$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing
+$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='builtin.func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing
 
 ===-------------------------------------------------------------------------===
                       ... Pass execution timing report ...
 ===-------------------------------------------------------------------------===
   Total Execution Time: 0.0249 seconds
 
    ---Wall Time---  --- Name ---
-   0.0058 ( 70.8%)  'func' Pipeline
+   0.0058 ( 70.8%)  'builtin.func' Pipeline
    0.0004 (  4.3%)    CSE
    0.0002 (  2.6%)      (A) DominanceInfo
    0.0004 (  4.8%)    VerifierPass
    0.0046 ( 55.4%)    Canonicalizer
    0.0005 (  6.2%)    VerifierPass
    0.0005 (  5.8%)  VerifierPass
    0.0014 ( 17.2%)  LLVMLoweringPass
    0.0005 (  6.2%)  VerifierPass
    0.0082 (100.0%)  Total
 ```
 
 ##### Multi-threaded Pass Timing
 
 When multi-threading is enabled in the pass manager the meaning of the display
 slightly changes. First, a new timing column is added, `User Time`, that
 displays the total time spent across all threads. Secondly, the `Wall Time`
 column displays the longest individual time spent amongst all of the threads.
 This means that the `Wall Time` column will continue to give an indicator on the
 perceived time, or clock time, whereas the `User Time` will display the total
 cpu time.
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse,canonicalize)' -convert-std-to-llvm -mlir-timing
 
 ===-------------------------------------------------------------------------===
                       ... Pass execution timing report ...
 ===-------------------------------------------------------------------------===
   Total Execution Time: 0.0078 seconds
 
    ---User Time---   ---Wall Time---  --- Name ---
-   0.0177 ( 88.5%)     0.0057 ( 71.3%)  'func' Pipeline
+   0.0177 ( 88.5%)     0.0057 ( 71.3%)  'builtin.func' Pipeline
    0.0044 ( 22.0%)     0.0015 ( 18.9%)    CSE
    0.0029 ( 14.5%)     0.0012 ( 15.2%)      (A) DominanceInfo
    0.0038 ( 18.9%)     0.0015 ( 18.7%)    VerifierPass
    0.0089 ( 44.6%)     0.0025 ( 31.1%)    Canonicalizer
    0.0006 (  3.0%)     0.0002 (  2.6%)    VerifierPass
    0.0004 (  2.2%)     0.0004 (  5.4%)  VerifierPass
    0.0013 (  6.5%)     0.0013 ( 16.3%)  LLVMLoweringPass
    0.0006 (  2.8%)     0.0006 (  7.0%)  VerifierPass

 this instrumentation:
 
 *   `print-ir-before=(comma-separated-pass-list)`
     *   Print the IR before each of the passes provided within the pass list.
 *   `print-ir-before-all`
     *   Print the IR before every pass in the pipeline.
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse)' -print-ir-before=cse
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse)' -print-ir-before=cse
 
 *** IR Dump Before CSE ***
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   %c1_i32_0 = constant 1 : i32
   return %c1_i32, %c1_i32_0 : i32, i32
 }
 ```
 
 *   `print-ir-after=(comma-separated-pass-list)`
     *   Print the IR after each of the passes provided within the pass list.
 *   `print-ir-after-all`
     *   Print the IR after every pass in the pipeline.
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse)' -print-ir-after=cse
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse)' -print-ir-after=cse
 
 *** IR Dump After CSE ***
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   return %c1_i32, %c1_i32 : i32, i32
 }
 ```
 
 *   `print-ir-after-change`
     *   Only print the IR after a pass if the pass mutated the IR. This helps to
         reduce the number of IR dumps for "uninteresting" passes.
     *   Note: Changes are detected by comparing a hash of the operation before
         and after the pass. This adds additional run-time to compute the hash of
         the IR, and in some rare cases may result in false-positives depending
         on the collision rate of the hash algorithm used.
     *   Note: This option should be used in unison with one of the other
         'print-ir-after' options above, as this option alone does not enable
         printing.
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse,cse)' -print-ir-after=cse -print-ir-after-change
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse,cse)' -print-ir-after=cse -print-ir-after-change
 
 *** IR Dump After CSE ***
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   return %c1_i32, %c1_i32 : i32, i32
 }
 ```
 
 *   `print-ir-after-failure`
     *   Only print IR after a pass failure.
     *   This option should *not* be used with the other `print-ir-after` flags
         above.
 
 ```shell
-$ mlir-opt foo.mlir -pass-pipeline='func(cse,bad-pass)' -print-ir-failure
+$ mlir-opt foo.mlir -pass-pipeline='builtin.func(cse,bad-pass)' -print-ir-failure
 
 *** IR Dump After BadPass Failed ***
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   return %c1_i32, %c1_i32 : i32, i32
 }
 ```
 
 *   `print-ir-module-scope`
     *   Always print the top-level module operation, regardless of pass type or
         operation nesting level.
     *   Note: Printing at module scope should only be used when multi-threading
         is disabled(`-mlir-disable-threading`)
 
 ```shell
-$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='func(cse)' -print-ir-after=cse -print-ir-module-scope
+$ mlir-opt foo.mlir -mlir-disable-threading -pass-pipeline='builtin.func(cse)' -print-ir-after=cse -print-ir-module-scope
 
-*** IR Dump After CSE ***  ('func' operation: @bar)
+*** IR Dump After CSE ***  ('builtin.func' operation: @bar)
 func @bar(%arg0: f32, %arg1: f32) -> f32 {
   ...
 }
 
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   %c1_i32_0 = constant 1 : i32
   return %c1_i32, %c1_i32_0 : i32, i32
 }
 
-*** IR Dump After CSE ***  ('func' operation: @simple_constant)
+*** IR Dump After CSE ***  ('builtin.func' operation: @simple_constant)
 func @bar(%arg0: f32, %arg1: f32) -> f32 {
   ...
 }
 
 func @simple_constant() -> (i32, i32) {
   %c1_i32 = constant 1 : i32
   return %c1_i32, %c1_i32 : i32, i32
 }
 ```
 
 ## Crash and Failure Reproduction
 
 The [pass manager](#pass-manager) in MLIR contains a builtin mechanism to
 generate reproducibles in the event of a crash, or a
 [pass failure](#pass-failure). This functionality can be enabled via
 `PassManager::enableCrashReproducerGeneration` or via the command line flag
 `pass-pipeline-crash-reproducer`. In either case, an argument is provided that
 corresponds to the output `.mlir` file name that the reproducible should be
 written to. The reproducible contains the configuration of the pass manager that
 was executing, as well as the initial IR before any passes were run. A potential
 reproducible may have the form:
 
 ```mlir
-// configuration: -pass-pipeline='func(cse,canonicalize),inline' -verify-each
+// configuration: -pass-pipeline='builtin.func(cse,canonicalize),inline' -verify-each
 
 module {
   func @foo() {
     ...
   }
 }
 ```
 

 
 Note: Local reproducer generation requires that multi-threading is
 disabled(`-mlir-disable-threading`)
 
 For example, if the failure in the previous example came from `canonicalize`,
 the following reproducer will be generated:
 
 ```mlir
-// configuration: -pass-pipeline='func(canonicalize)' -verify-each -mlir-disable-threading
+// configuration: -pass-pipeline='builtin.func(canonicalize)' -verify-each -mlir-disable-threading
 
 module {
   func @foo() {
     ...
   }
 }
 ```

mlir/docs/SymbolsAndSymbolTables.md

 
 ## Symbol
 
 A `Symbol` is a named operation that resides immediately within a region that
 defines a [`SymbolTable`](#symbol-table). The name of a symbol *must* be unique
 within the parent `SymbolTable`. This name is semantically similarly to an SSA
 result value, and may be referred to by other operations to provide a symbolic
 link, or use, to the symbol. An example of a `Symbol` operation is
-[`func`](Dialects/Builtin.md/#func-mlirfuncop). `func` defines a symbol name, which is
-[referred to](#referencing-a-symbol) by operations like
+[`builtin.func`](Dialects/Builtin.md/#func-mlirfuncop). `builtin.func` defines a
+symbol name, which is [referred to](#referencing-a-symbol) by operations like
 [`std.call`](Dialects/Standard.md/#stdcall-callop).
 
 ### Defining or declaring a Symbol
 
 A `Symbol` operation should use the `SymbolOpInterface` interface to provide the
-necessary verification and accessors; it also supports
-operations, such as `module`, that conditionally define a symbol. `Symbol`s must
-have the following properties:
+necessary verification and accessors; it also supports operations, such as
+`builtin.module`, that conditionally define a symbol. `Symbol`s must have the
+following properties:
 
 *   A `StringAttr` attribute named
     'SymbolTable::getSymbolAttrName()'(`sym_name`).
     -   This attribute defines the symbolic 'name' of the operation.
 *   An optional `StringAttr` attribute named
     'SymbolTable::getVisibilityAttrName()'(`sym_visibility`)
     -   This attribute defines the [visibility](#symbol-visibility) of the
         symbol, or more specifically in-which scopes it may be accessed.

 within a symbol table. It may optionally contain a set of nested references that
 further resolve to a symbol nested within a different symbol table. When
 resolving a nested reference, each non-leaf reference must refer to a symbol
 operation that is also a [symbol table](#symbol-table).
 
 Below is an example of how an operation can reference a symbol operation:
 
 ```mlir
-// This `func` operation defines a symbol named `symbol`.
+// This `builtin.func` operation defines a symbol named `symbol`.
 func @symbol()
 
 // Our `foo.user` operation contains a SymbolRefAttr with the name of the
 // `symbol` func.
 "foo.user"() {uses = [@symbol]} : () -> ()
 
 // Symbol references resolve to the nearest parent operation that defines a
 // symbol table, so we can have references with arbitrary nesting levels.
module {
   // Error. We resolve references with respect to the closest parent operation
   // that defines a symbol table, so this reference can't be resolved.
   "foo.user"() {uses = [@symbol]} : () -> ()
 }
 
 // Here we define another nested symbol table, except this time it also defines
 // a symbol.
 module @module_symbol {
-  // This `func` operation defines a symbol named `nested_symbol`.
+  // This `builtin.func` operation defines a symbol named `nested_symbol`.
   func @nested_symbol()
 }
 
 // Our `foo.user` operation may refer to the nested symbol, by resolving through
 // the parent.
 "foo.user"() {uses = [@module_symbol::@nested_symbol]} : () -> ()
 ```
 

mlir/docs/Tutorials/UnderstandingTheIRStructure.md

 
 ### Example
 
 The Pass introduced in the previous section can be applied on the following IR
 with `mlir-opt -test-print-nesting -allow-unregistered-dialect
 llvm-project/mlir/test/IR/print-ir-nesting.mlir`:
 
 ```mlir
-"module"() ( {
+"builtin.module"() ( {
   %0:4 = "dialect.op1"() {"attribute name" = 42 : i32} : () -> (i1, i16, i32, i64)
   "dialect.op2"() ( {
     "dialect.innerop1"(%0#0, %0#1) : (i1, i16) -> ()
   },  {
     "dialect.innerop2"() : () -> ()
     "dialect.innerop3"(%0#0, %0#2, %0#3)[^bb1, ^bb2] : (i1, i32, i64) -> ()
   ^bb1(%1: i32):  // pred: ^bb0
     "dialect.innerop4"() : () -> ()
     "dialect.innerop5"() : () -> ()
   ^bb2(%2: i64):  // pred: ^bb0
     "dialect.innerop6"() : () -> ()
     "dialect.innerop7"() : () -> ()
   }) {"other attribute" = 42 : i64} : () -> ()
 }) : () -> ()
 ```
 
 And will yield the following output:
 
 ```
-visiting op: 'module' with 0 operands and 0 results
+visiting op: 'builtin.module' with 0 operands and 0 results
  1 nested regions:
   Region with 1 blocks:
     Block with 0 arguments, 0 successors, and 3 operations
       visiting op: 'dialect.op1' with 0 operands and 4 results
       1 attributes:
        - 'attribute name' : '42 : i32'
        0 nested regions:
       visiting op: 'dialect.op2' with 0 operands and 0 results

mlir/include/mlir/Dialect/PDL/IR/PDLOps.td

let description = [{
     1) A single !pdl.range<type>:
 
     In this case, the single range is treated as all of the result types of the
     operation.
 
     ```mlir
     // Define an instance with single range of types.
     %allResultTypes = pdl.types
-    %op = pdl.operation "unrealized_conversion_cast" -> (%allResultTypes : !pdl.types)
+    %op = pdl.operation "builtin.unrealized_conversion_cast" -> (%allResultTypes : !pdl.types)
     ```
 
     2) A variadic number of either !pdl.type or !pdl.range<type>:
 
     In this case, the inputs are expected to correspond with the result groups
     defined on the operation in ODS.
 
     ```tablgen

mlir/include/mlir/IR/BuiltinDialect.td

 #define BUILTIN_BASE
 
 include "mlir/IR/OpBase.td"
 
 def Builtin_Dialect : Dialect {
   let summary =
     "A dialect containing the builtin Attributes, Operations, and Types";
 
-  let name = "";
+  let name = "builtin";
   let cppNamespace = "::mlir";
   let extraClassDeclaration = [{
   private:
     // Register the builtin Attributes.
     void registerAttributes();
     // Register the builtin Location Attributes.
     void registerLocationAttributes();
     // Register the builtin Types.
     void registerTypes();
 
   public:
   }];
 }
 
 #endif // BUILTIN_BASE

mlir/include/mlir/IR/Dialect.h

SmallVector<std::pair<TypeID, ObjectInterfaceAllocatorFunction>, 2>
         objectInterfaces;
   };
 
   using MapTy =
       std::map<std::string, std::pair<TypeID, DialectAllocatorFunction>>;
   using InterfaceMapTy = DenseMap<TypeID, DelayedInterfaces>;
 
 public:
-  explicit DialectRegistry() {}
+  explicit DialectRegistry();
 
   template <typename ConcreteDialect>
   void insert() {
     insert(TypeID::get<ConcreteDialect>(),
            ConcreteDialect::getDialectNamespace(),
            static_cast<DialectAllocatorFunction>(([](MLIRContext *ctx) {
              // Just allocate the dialect, the context
              // takes ownership of it.
public:
 
   /// Add an external op interface model for an op that belongs to a dialect,
   /// both provided as template parameters. The dialect must be present in the
   /// registry.
   template <typename OpTy, typename ModelTy>
   void addOpInterface() {
     StringRef opName = OpTy::getOperationName();
     StringRef dialectName = opName.split('.').first;
-    addObjectInterface(dialectName == opName ? "" : dialectName,
-                       ModelTy::Interface::getInterfaceID(),
+    addObjectInterface(dialectName, ModelTy::Interface::getInterfaceID(),
                        [](MLIRContext *context) {
                          OpTy::template attachInterface<ModelTy>(*context);
                        });
   }
 
   /// Add an external attribute interface model for an attribute type `AttrTy`
   /// that is going to belong to `DialectTy`. The dialect must be present in the
   /// registry.

mlir/include/mlir/Pass/PassManager.h

 
 /// The main pass manager and pipeline builder.
 class PassManager : public OpPassManager {
 public:
   /// Create a new pass manager under the given context with a specific nesting
   /// style. The created pass manager can schedule operations that match
   /// `operationName`.
   PassManager(MLIRContext *ctx, Nesting nesting = Nesting::Explicit,
-              StringRef operationName = "module");
+              StringRef operationName = "builtin.module");
   PassManager(MLIRContext *ctx, StringRef operationName)
       : PassManager(ctx, Nesting::Explicit, operationName) {}
   ~PassManager();
 
   /// Run the passes within this manager on the provided operation. The
   /// specified operation must have the same name as the one provided the pass
   /// manager on construction.
   LogicalResult run(Operation *op);

mlir/include/mlir/Transforms/LoopUtils.h

 ///    scf.for %i = %lb + %step * (threadIdx.x + blockIdx.x * blockDim.x)
 ///       to %ub step %gridDim.x * blockDim.x * %step {
 ///      ...
 ///    }
 /// ```
 void mapLoopToProcessorIds(scf::ForOp forOp, ArrayRef<Value> processorId,
                            ArrayRef<Value> numProcessors);
 
-/// Gathers all AffineForOps in 'func' grouped by loop depth.
+/// Gathers all AffineForOps in 'builtin.func' grouped by loop depth.
 void gatherLoops(FuncOp func,
                  std::vector<SmallVector<AffineForOp, 2>> &depthToLoops);
 
 /// Creates an AffineForOp while ensuring that the lower and upper bounds are
 /// canonicalized, i.e., unused and duplicate operands are removed, any constant
 /// operands propagated/folded in, and duplicate bound maps dropped.
 AffineForOp createCanonicalizedAffineForOp(OpBuilder b, Location loc,
                                            ValueRange lbOperands,

mlir/lib/Bindings/Python/IRCore.cpp

 }
 
 //------------------------------------------------------------------------------
 // PyDialect, PyDialectDescriptor, PyDialects
 //------------------------------------------------------------------------------
 
 MlirDialect PyDialects::getDialectForKey(const std::string &key,
                                          bool attrError) {
-  // If the "std" dialect was asked for, substitute the empty namespace :(
-  static const std::string emptyKey;
-  const std::string *canonKey = key == "std" ? &emptyKey : &key;
-  MlirDialect dialect = mlirContextGetOrLoadDialect(
-      getContext()->get(), {canonKey->data(), canonKey->size()});
+  MlirDialect dialect = mlirContextGetOrLoadDialect(getContext()->get(),
+                                                    {key.data(), key.size()});
   if (mlirDialectIsNull(dialect)) {
     throw SetPyError(attrError ? PyExc_AttributeError : PyExc_IndexError,
                      Twine("Dialect '") + key + "' not found");
   }
   return dialect;
 }
 
 //------------------------------------------------------------------------------

mlir/lib/Dialect/Linalg/Transforms/ComprehensiveBufferize.cpp

void LinalgComprehensiveModuleBufferize::runOnOperation() {
   // Post-pass cleanup of inplaceable and buffer_layout attributes.
   moduleOp.walk(
       [&](Operation *op) { op->removeAttr(kInPlaceResultsAttrName); });
   moduleOp.walk([&](FuncOp op) {
     for (BlockArgument bbArg : op.getArguments())
       removeBufferizationFuncArguments(bbArg);
   });
 
-  OpPassManager cleanupPipeline(OpPassManager("module"));
+  OpPassManager cleanupPipeline("builtin.module");
   cleanupPipeline.addPass(createCanonicalizerPass());
   cleanupPipeline.addPass(createCSEPass());
   cleanupPipeline.addPass(createLoopInvariantCodeMotionPass());
   (void)runPipeline(cleanupPipeline, moduleOp);
 }
 
 std::unique_ptr<Pass> mlir::createLinalgComprehensiveModuleBufferizePass() {
   return std::make_unique<LinalgComprehensiveModuleBufferize>();
 }

mlir/lib/IR/BuiltinDialect.cpp

 }
 
 //===----------------------------------------------------------------------===//
 // FuncOp
 //===----------------------------------------------------------------------===//
 
 FuncOp FuncOp::create(Location location, StringRef name, FunctionType type,
                       ArrayRef<NamedAttribute> attrs) {
-  OperationState state(location, "func");
   OpBuilder builder(location->getContext());
+  OperationState state(location, getOperationName());
   FuncOp::build(builder, state, name, type, attrs);
   return cast<FuncOp>(Operation::create(state));
 }
 FuncOp FuncOp::create(Location location, StringRef name, FunctionType type,
                       Operation::dialect_attr_range attrs) {
   SmallVector<NamedAttribute, 8> attrRef(attrs);
   return create(location, name, type, llvm::makeArrayRef(attrRef));
 }

mlir/lib/IR/BuiltinTypes.cpp

 
 /// Return true if the specified element type is ok in a tensor.
 bool TensorType::isValidElementType(Type type) {
   // Note: Non standard/builtin types are allowed to exist within tensor
   // types. Dialects are expected to verify that tensor types have a valid
   // element type within that dialect.
   return type.isa<ComplexType, FloatType, IntegerType, OpaqueType, VectorType,
                   IndexType>() ||
-         !type.getDialect().getNamespace().empty();
+         !llvm::isa<BuiltinDialect>(type.getDialect());
 }
 
 //===----------------------------------------------------------------------===//
 // RankedTensorType
 //===----------------------------------------------------------------------===//
 
 LogicalResult
 RankedTensorType::verify(function_ref<InFlightDiagnostic()> emitError,

mlir/lib/IR/Dialect.cpp

 using namespace detail;
 
 DialectAsmParser::~DialectAsmParser() {}
 
 //===----------------------------------------------------------------------===//
 // DialectRegistry
 //===----------------------------------------------------------------------===//
 
+DialectRegistry::DialectRegistry() { insert<BuiltinDialect>(); }
+
 void DialectRegistry::addDialectInterface(
     StringRef dialectName, TypeID interfaceTypeID,
     DialectInterfaceAllocatorFunction allocator) {
   assert(allocator && "unexpected null interface allocation function");
   auto it = registry.find(dialectName.str());
   assert(it != registry.end() &&
          "adding an interface for an unregistered dialect");
 
void DialectRegistry::addDialectInterface(
 
   ifaces.dialectInterfaces.emplace_back(interfaceTypeID, allocator);
 }
 
 void DialectRegistry::addObjectInterface(
     StringRef dialectName, TypeID interfaceTypeID,
     ObjectInterfaceAllocatorFunction allocator) {
   assert(allocator && "unexpected null interface allocation function");
-
-  // Builtin dialect has an empty prefix and is always registered.
-  TypeID dialectTypeID;
-  if (!dialectName.empty()) {
-    auto it = registry.find(dialectName.str());
-    assert(it != registry.end() &&
-           "adding an interface for an op from an unregistered dialect");
-    dialectTypeID = it->second.first;
-  } else {
-    dialectTypeID = TypeID::get<BuiltinDialect>();
-  }
 
-  auto &ifaces = interfaces[dialectTypeID];
+  auto &ifaces = interfaces[it->second.first];
   for (const auto &kvp : ifaces.objectInterfaces) {
     if (kvp.first == interfaceTypeID) {
       LLVM_DEBUG(llvm::dbgs()
                  << "[" DEBUG_TYPE
                     "] repeated interface object interface registration");
       return;
     }
   }
LogicalResult Dialect::printOperation(Operation *op,
   assert(op->getDialect() == this &&
          "Dialect hook invoked on non-dialect owned operation");
   return failure();
 }
 
 /// Utility function that returns if the given string is a valid dialect
 /// namespace.
 bool Dialect::isValidNamespace(StringRef str) {
-  if (str.empty())
-    return true;
   llvm::Regex dialectNameRegex("^[a-zA-Z_][a-zA-Z_0-9\\$]*$");
   return dialectNameRegex.match(str);
 }
 
 /// Register a set of dialect interfaces with this dialect instance.
 void Dialect::addInterface(std::unique_ptr<DialectInterface> interface) {
   auto it = registeredInterfaces.try_emplace(interface->getID(),
                                              std::move(interface));

mlir/lib/IR/Operation.cpp

OperationName::OperationName(StringRef name, MLIRContext *context) {
   else
     representation = Identifier::get(name, context);
 }
 
 /// Return the name of the dialect this operation is registered to.
 StringRef OperationName::getDialectNamespace() const {
   if (Dialect *dialect = getDialect())
     return dialect->getNamespace();
-  return representation.get<Identifier>().strref().split('.').first;
+  return getStringRef().split('.').first;
 }
 
 /// Return the operation name with dialect name stripped, if it has one.
 StringRef OperationName::stripDialect() const {
-  auto splitName = getStringRef().split(".");
-  return splitName.second.empty() ? splitName.first : splitName.second;
+  return getStringRef().split('.').second;
 }
 
 /// Return the name of this operation. This always succeeds.
 StringRef OperationName::getStringRef() const {
   return getIdentifier().strref();
 }
 
 /// Return the name of this operation as an identifier. This always succeeds.

mlir/lib/Parser/Parser.cpp

 /// Create and remember a new placeholder for a forward reference.
 Value OperationParser::createForwardRefPlaceholder(SMLoc loc, Type type) {
   // Forward references are always created as operations, because we just need
   // something with a def/use chain.
   //
   // We create these placeholders as having an empty name, which we know
   // cannot be created through normal user input, allowing us to distinguish
   // them.
-  auto name = OperationName("unrealized_conversion_cast", getContext());
+  auto name = OperationName("builtin.unrealized_conversion_cast", getContext());
   auto *op = Operation::create(
       getEncodedSourceLocation(loc), name, type, /*operands=*/{},
       /*attributes=*/llvm::None, /*successors=*/{}, /*numRegions=*/0);
   forwardRefPlaceholders[op->getResult(0)] = loc;
   return op->getResult(0);
 }
 
 //===----------------------------------------------------------------------===//
if (opDefinition) {
     if (opName.contains('.')) {
       // This op has a dialect, we try to check if we can register it in the
       // context on the fly.
       StringRef dialectName = opName.split('.').first;
       dialect = getContext()->getLoadedDialect(dialectName);
       if (!dialect && (dialect = getContext()->getOrLoadDialect(dialectName)))
         opDefinition = AbstractOperation::lookup(opName, getContext());
     } else {
-      // If the operation name has no namespace prefix we treat it as a standard
-      // operation and prefix it with "std".
-      // TODO: Would it be better to just build a mapping of the registered
-      // operations in the standard dialect?
-      if (getContext()->getOrLoadDialect("std")) {
+      // If the operation name has no namespace prefix we treat it as a builtin
+      // or standard operation and prefix it with "builtin" or "std".
+      // TODO: Remove the special casing here.
+      opDefinition = AbstractOperation::lookup(Twine("builtin." + opName).str(),
+                                               getContext());
+      if (!opDefinition && getContext()->getOrLoadDialect("std")) {
         opDefinition = AbstractOperation::lookup(Twine("std." + opName).str(),
                                                  getContext());
-        if (opDefinition)
-          opName = opDefinition->name.strref();
-      }
-    }
-  }
+      }
 
   // This is the actual hook for the custom op parsing, usually implemented by
   // the op itself (`Op::parse()`). We retrieve it either from the
   // AbstractOperation or from the Dialect.
   function_ref<ParseResult(OpAsmParser &, OperationState &)> parseAssemblyFn;
   bool isIsolatedFromAbove = false;
 
   if (opDefinition) {

mlir/lib/Pass/PassRegistry.cpp

private:
   /// parameter corresponds to the raw location within the pipeline string. This
   /// should always return failure.
   using ErrorHandlerT = function_ref<LogicalResult(const char *, Twine)>;
 
   /// A struct to capture parsed pass pipeline names.
   ///
   /// A pipeline is defined as a series of names, each of which may in itself
   /// recursively contain a nested pipeline. A name is either the name of a pass
-  /// (e.g. "cse") or the name of an operation type (e.g. "func"). If the name
-  /// is the name of a pass, the InnerPipeline is empty, since passes cannot
-  /// contain inner pipelines.
+  /// (e.g. "cse") or the name of an operation type (e.g. "builtin.func"). If
+  /// the name is the name of a pass, the InnerPipeline is empty, since passes
+  /// cannot contain inner pipelines.
   struct PipelineElement {
     PipelineElement(StringRef name) : name(name), registryEntry(nullptr) {}
 
     StringRef name;
     StringRef options;
     const PassRegistryEntry *registryEntry;
     std::vector<PipelineElement> innerPipeline;
   };

mlir/lib/Transforms/Utils/LoopUtils.cpp

gatherLoopsInBlock(Block *block, unsigned currLoopDepth,
   for (auto &op : *block) {
     if (auto forOp = dyn_cast<AffineForOp>(op)) {
       depthToLoops[currLoopDepth].push_back(forOp);
       gatherLoopsInBlock(forOp.getBody(), currLoopDepth + 1, depthToLoops);
     }
   }
 }
 
-/// Gathers all AffineForOps in 'func' grouped by loop depth.
+/// Gathers all AffineForOps in 'builtin.func' grouped by loop depth.
 void mlir::gatherLoops(FuncOp func,
                        std::vector<SmallVector<AffineForOp, 2>> &depthToLoops) {
   for (auto &block : func)
     gatherLoopsInBlock(&block, /*currLoopDepth=*/0, depthToLoops);
 
   // Remove last loop level from output since it's empty.
   if (!depthToLoops.empty()) {
     assert(depthToLoops.back().empty() && "Last loop level is not empty?");

mlir/test/Analysis/test-alias-analysis-modref.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(test-alias-analysis-modref)' -split-input-file -allow-unregistered-dialect 2>&1 | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(test-alias-analysis-modref)' -split-input-file -allow-unregistered-dialect 2>&1 | FileCheck %s
 
 // CHECK-LABEL: Testing : "no_side_effects"
 // CHECK: alloc -> func.region0#0: NoModRef
 // CHECK: dealloc -> func.region0#0: NoModRef
 // CHECK: return -> func.region0#0: NoModRef
 func @no_side_effects(%arg: memref<2xf32>) attributes {test.ptr = "func"} {
   %1 = memref.alloc() {test.ptr = "alloc"} : memref<8x64xf32>
   memref.dealloc %1 {test.ptr = "dealloc"} : memref<8x64xf32>

mlir/test/Analysis/test-alias-analysis.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(test-alias-analysis)' -split-input-file -allow-unregistered-dialect 2>&1 | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(test-alias-analysis)' -split-input-file -allow-unregistered-dialect 2>&1 | FileCheck %s
 
 // CHECK-LABEL: Testing : "simple"
 // CHECK-DAG: func.region0#0 <-> func.region0#1: MayAlias
 
 // CHECK-DAG: alloca_1#0 <-> alloca_2#0: NoAlias
 // CHECK-DAG: alloca_1#0 <-> alloc_1#0: NoAlias
 // CHECK-DAG: alloca_1#0 <-> alloc_2#0: NoAlias
 // CHECK-DAG: alloca_1#0 <-> func.region0#0: NoAlias

mlir/test/CAPI/ir.c

MlirAttribute funcNameAttr =
       mlirAttributeParseGet(ctx, mlirStringRefCreateFromCString("\"add\""));
   MlirNamedAttribute funcAttrs[] = {
       mlirNamedAttributeGet(
           mlirIdentifierGet(ctx, mlirStringRefCreateFromCString("type")),
           funcTypeAttr),
       mlirNamedAttributeGet(
           mlirIdentifierGet(ctx, mlirStringRefCreateFromCString("sym_name")),
           funcNameAttr)};
-  MlirOperationState funcState =
-      mlirOperationStateGet(mlirStringRefCreateFromCString("func"), location);
+  MlirOperationState funcState = mlirOperationStateGet(
+      mlirStringRefCreateFromCString("builtin.func"), location);
   mlirOperationStateAddAttributes(&funcState, 2, funcAttrs);
   mlirOperationStateAddOwnedRegions(&funcState, 1, &funcBodyRegion);
   MlirOperation func = mlirOperationCreate(&funcState);
   mlirBlockInsertOwnedOperation(moduleBody, 0, func);
 
   MlirType indexType =
       mlirTypeParseGet(ctx, mlirStringRefCreateFromCString("index"));
   MlirAttribute indexZeroLiteral =

mlir/test/CAPI/pass.c

"}"));
   // clang-format on
   if (mlirModuleIsNull(module)) {
     fprintf(stderr, "Unexpected failure parsing module.\n");
     exit(EXIT_FAILURE);
   }
 
   // Run the print-op-stats pass on the top-level module:
   // CHECK-LABEL: Operations encountered:
-  // CHECK: func              , 1
+  // CHECK: builtin.func      , 1
   // CHECK: std.addi          , 1
   // CHECK: std.return        , 1
   {
     MlirPassManager pm = mlirPassManagerCreate(ctx);
     MlirPass printOpStatPass = mlirCreateTransformsPrintOpStats();
     mlirPassManagerAddOwnedPass(pm, printOpStatPass);
     MlirLogicalResult success = mlirPassManagerRun(pm, module);
     if (mlirLogicalResultIsFailure(success)) {

 "  }                                                                        \n"
 "}"));
   // clang-format on
   if (mlirModuleIsNull(module))
     exit(1);
 
   // Run the print-op-stats pass on functions under the top-level module:
   // CHECK-LABEL: Operations encountered:
-  // CHECK: func              , 1
+  // CHECK: builtin.func      , 1
   // CHECK: std.addi          , 1
   // CHECK: std.return        , 1
   {
     MlirPassManager pm = mlirPassManagerCreate(ctx);
     MlirOpPassManager nestedFuncPm = mlirPassManagerGetNestedUnder(
-        pm, mlirStringRefCreateFromCString("func"));
+        pm, mlirStringRefCreateFromCString("builtin.func"));
     MlirPass printOpStatPass = mlirCreateTransformsPrintOpStats();
     mlirOpPassManagerAddOwnedPass(nestedFuncPm, printOpStatPass);
     MlirLogicalResult success = mlirPassManagerRun(pm, module);
     if (mlirLogicalResultIsFailure(success))
       exit(2);
     mlirPassManagerDestroy(pm);
   }
   // Run the print-op-stats pass on functions under the nested module:
   // CHECK-LABEL: Operations encountered:
-  // CHECK: func              , 1
+  // CHECK: builtin.func      , 1
   // CHECK: std.addf          , 1
   // CHECK: std.return        , 1
   {
     MlirPassManager pm = mlirPassManagerCreate(ctx);
     MlirOpPassManager nestedModulePm = mlirPassManagerGetNestedUnder(
-        pm, mlirStringRefCreateFromCString("module"));
+        pm, mlirStringRefCreateFromCString("builtin.module"));
     MlirOpPassManager nestedFuncPm = mlirOpPassManagerGetNestedUnder(
-        nestedModulePm, mlirStringRefCreateFromCString("func"));
+        nestedModulePm, mlirStringRefCreateFromCString("builtin.func"));
     MlirPass printOpStatPass = mlirCreateTransformsPrintOpStats();
     mlirOpPassManagerAddOwnedPass(nestedFuncPm, printOpStatPass);
     MlirLogicalResult success = mlirPassManagerRun(pm, module);
     if (mlirLogicalResultIsFailure(success))
       exit(2);
     mlirPassManagerDestroy(pm);
   }
 
   mlirModuleDestroy(module);
   mlirContextDestroy(ctx);
 }
 
 static void printToStderr(MlirStringRef str, void *userData) {
   (void)userData;
   fwrite(str.data, 1, str.length, stderr);
 }
 
 void testPrintPassPipeline() {
   MlirContext ctx = mlirContextCreate();
   MlirPassManager pm = mlirPassManagerCreate(ctx);
   // Populate the pass-manager
   MlirOpPassManager nestedModulePm = mlirPassManagerGetNestedUnder(
-      pm, mlirStringRefCreateFromCString("module"));
+      pm, mlirStringRefCreateFromCString("builtin.module"));
   MlirOpPassManager nestedFuncPm = mlirOpPassManagerGetNestedUnder(
-      nestedModulePm, mlirStringRefCreateFromCString("func"));
+      nestedModulePm, mlirStringRefCreateFromCString("builtin.func"));
   MlirPass printOpStatPass = mlirCreateTransformsPrintOpStats();
   mlirOpPassManagerAddOwnedPass(nestedFuncPm, printOpStatPass);
 
   // Print the top level pass manager
-  // CHECK: Top-level: module(func(print-op-stats))
+  // CHECK: Top-level: builtin.module(builtin.func(print-op-stats))
   fprintf(stderr, "Top-level: ");
   mlirPrintPassPipeline(mlirPassManagerGetAsOpPassManager(pm), printToStderr,
                         NULL);
   fprintf(stderr, "\n");
 
   // Print the pipeline nested one level down
-  // CHECK: Nested Module: func(print-op-stats)
+  // CHECK: Nested Module: builtin.func(print-op-stats)
   fprintf(stderr, "Nested Module: ");
   mlirPrintPassPipeline(nestedModulePm, printToStderr, NULL);
   fprintf(stderr, "\n");
 
   // Print the pipeline nested two levels down
   // CHECK: Nested Module>Func: print-op-stats
   fprintf(stderr, "Nested Module>Func: ");
   mlirPrintPassPipeline(nestedFuncPm, printToStderr, NULL);
   fprintf(stderr, "\n");
 
   mlirPassManagerDestroy(pm);
   mlirContextDestroy(ctx);
 }
 
 void testParsePassPipeline() {
   MlirContext ctx = mlirContextCreate();
   MlirPassManager pm = mlirPassManagerCreate(ctx);
   // Try parse a pipeline.
   MlirLogicalResult status = mlirParsePassPipeline(
       mlirPassManagerGetAsOpPassManager(pm),
       mlirStringRefCreateFromCString(
-          "module(func(print-op-stats), func(print-op-stats))"));
+          "builtin.module(builtin.func(print-op-stats), builtin.func(print-op-stats))"));
   // Expect a failure, we haven't registered the print-op-stats pass yet.
   if (mlirLogicalResultIsSuccess(status)) {
     fprintf(stderr, "Unexpected success parsing pipeline without registering the pass\n");
     exit(EXIT_FAILURE);
   }
   // Try again after registrating the pass.
   mlirRegisterTransformsPrintOpStats();
   status = mlirParsePassPipeline(
       mlirPassManagerGetAsOpPassManager(pm),
       mlirStringRefCreateFromCString(
-          "module(func(print-op-stats), func(print-op-stats))"));
+          "builtin.module(builtin.func(print-op-stats), builtin.func(print-op-stats))"));
   // Expect a failure, we haven't registered the print-op-stats pass yet.
   if (mlirLogicalResultIsFailure(status)) {
     fprintf(stderr, "Unexpected failure parsing pipeline after registering the pass\n");
     exit(EXIT_FAILURE);
   }
 
-  // CHECK: Round-trip: module(func(print-op-stats), func(print-op-stats))
+  // CHECK: Round-trip: builtin.module(builtin.func(print-op-stats), builtin.func(print-op-stats))
   fprintf(stderr, "Round-trip: ");
   mlirPrintPassPipeline(mlirPassManagerGetAsOpPassManager(pm), printToStderr,
                         NULL);
   fprintf(stderr, "\n");
 }
 
 int main() {
   testRunPassOnModule();
   testRunPassOnNestedModule();
   testPrintPassPipeline();
   testParsePassPipeline();
   return 0;
 }

mlir/test/Conversion/AsyncToLLVM/convert-runtime-to-llvm.mlir

func @create_value() {
   // CHECK: %[[VALUE:.*]] = call @mlirAsyncRuntimeCreateValue(%[[SIZE]])
   %0 = async.runtime.create : !async.value<f32>
   return
 }
 
 // CHECK-LABEL: @create_group
 func @create_group() {
   // CHECK: %[[C:.*]] = constant 1 : index
-  // CHECK: %[[S:.*]] = unrealized_conversion_cast %[[C]] : index to i64
+  // CHECK: %[[S:.*]] = builtin.unrealized_conversion_cast %[[C]] : index to i64
   %c = constant 1 : index
   // CHECK: %[[GROUP:.*]] = call @mlirAsyncRuntimeCreateGroup(%[[S]])
   %0 = async.runtime.create_group  %c: !async.group
   return
 }
 
 // CHECK-LABEL: @set_token_available
 func @set_token_available() {

mlir/test/Conversion/ComplexToLLVM/convert-to-llvm.mlir

 // RUN: mlir-opt %s -convert-complex-to-llvm | FileCheck %s
 
 // CHECK-LABEL: func @complex_create
 // CHECK-SAME:    (%[[REAL0:.*]]: f32, %[[IMAG0:.*]]: f32)
 // CHECK-NEXT:    %[[CPLX0:.*]] = llvm.mlir.undef : !llvm.struct<(f32, f32)>
 // CHECK-NEXT:    %[[CPLX1:.*]] = llvm.insertvalue %[[REAL0]], %[[CPLX0]][0] : !llvm.struct<(f32, f32)>
 // CHECK-NEXT:    %[[CPLX2:.*]] = llvm.insertvalue %[[IMAG0]], %[[CPLX1]][1] : !llvm.struct<(f32, f32)>
 func @complex_create(%real: f32, %imag: f32) -> complex<f32> {
   %cplx2 = complex.create %real, %imag : complex<f32>
   return %cplx2 : complex<f32>
 }
 
 // CHECK-LABEL: func @complex_extract
 // CHECK-SAME:    (%[[CPLX:.*]]: complex<f32>)
-// CHECK-NEXT:    %[[CAST0:.*]] = unrealized_conversion_cast %[[CPLX]] : complex<f32> to !llvm.struct<(f32, f32)>
+// CHECK-NEXT:    %[[CAST0:.*]] = builtin.unrealized_conversion_cast %[[CPLX]] : complex<f32> to !llvm.struct<(f32, f32)>
 // CHECK-NEXT:    %[[REAL:.*]] = llvm.extractvalue %[[CAST0]][0] : !llvm.struct<(f32, f32)>
-// CHECK-NEXT:    %[[CAST1:.*]] = unrealized_conversion_cast %[[CPLX]] : complex<f32> to !llvm.struct<(f32, f32)>
+// CHECK-NEXT:    %[[CAST1:.*]] = builtin.unrealized_conversion_cast %[[CPLX]] : complex<f32> to !llvm.struct<(f32, f32)>
 // CHECK-NEXT:    %[[IMAG:.*]] = llvm.extractvalue %[[CAST1]][1] : !llvm.struct<(f32, f32)>
 func @complex_extract(%cplx: complex<f32>) {
   %real1 = complex.re %cplx : complex<f32>
   %imag1 = complex.im %cplx : complex<f32>
   return
 }
 
 // CHECK-LABEL: func @complex_addition

 }
 
 // CHECK-LABEL: func @complex_div
 // CHECK-SAME:    %[[LHS:.*]]: complex<f32>, %[[RHS:.*]]: complex<f32>
 func @complex_div(%lhs: complex<f32>, %rhs: complex<f32>) -> complex<f32> {
   %div = complex.div %lhs, %rhs : complex<f32>
   return %div : complex<f32>
 }
-// CHECK: %[[CASTED_LHS:.*]] = unrealized_conversion_cast %[[LHS]] : complex<f32> to ![[C_TY:.*>]]
-// CHECK: %[[CASTED_RHS:.*]] = unrealized_conversion_cast %[[RHS]] : complex<f32> to ![[C_TY]]
+// CHECK: %[[CASTED_LHS:.*]] = builtin.unrealized_conversion_cast %[[LHS]] : complex<f32> to ![[C_TY:.*>]]
+// CHECK: %[[CASTED_RHS:.*]] = builtin.unrealized_conversion_cast %[[RHS]] : complex<f32> to ![[C_TY]]
 
 // CHECK: %[[LHS_RE:.*]] = llvm.extractvalue %[[CASTED_LHS]][0] : ![[C_TY]]
 // CHECK: %[[LHS_IM:.*]] = llvm.extractvalue %[[CASTED_LHS]][1] : ![[C_TY]]
 // CHECK: %[[RHS_RE:.*]] = llvm.extractvalue %[[CASTED_RHS]][0] : ![[C_TY]]
 // CHECK: %[[RHS_IM:.*]] = llvm.extractvalue %[[CASTED_RHS]][1] : ![[C_TY]]
 
 // CHECK: %[[RESULT_0:.*]] = llvm.mlir.undef : ![[C_TY]]
 

 // CHECK-DAG: %[[IMAG_TMP_1:.*]] = llvm.fmul %[[LHS_RE]], %[[RHS_IM]]  : f32
 // CHECK: %[[IMAG_TMP_2:.*]] = llvm.fsub %[[IMAG_TMP_0]], %[[IMAG_TMP_1]]  : f32
 
 // CHECK: %[[REAL:.*]] = llvm.fdiv %[[REAL_TMP_2]], %[[SQ_NORM]]  : f32
 // CHECK: %[[RESULT_1:.*]] = llvm.insertvalue %[[REAL]], %[[RESULT_0]][0] : ![[C_TY]]
 // CHECK: %[[IMAG:.*]] = llvm.fdiv %[[IMAG_TMP_2]], %[[SQ_NORM]]  : f32
 // CHECK: %[[RESULT_2:.*]] = llvm.insertvalue %[[IMAG]], %[[RESULT_1]][1] : ![[C_TY]]
 //
-// CHECK: %[[CASTED_RESULT:.*]] = unrealized_conversion_cast %[[RESULT_2]] : ![[C_TY]] to complex<f32>
+// CHECK: %[[CASTED_RESULT:.*]] = builtin.unrealized_conversion_cast %[[RESULT_2]] : ![[C_TY]] to complex<f32>
 // CHECK: return %[[CASTED_RESULT]] : complex<f32>
 
 // CHECK-LABEL: func @complex_mul
 // CHECK-SAME:    %[[LHS:.*]]: complex<f32>, %[[RHS:.*]]: complex<f32>
 func @complex_mul(%lhs: complex<f32>, %rhs: complex<f32>) -> complex<f32> {
   %mul = complex.mul %lhs, %rhs : complex<f32>
   return %mul : complex<f32>
 }
-// CHECK: %[[CASTED_LHS:.*]] = unrealized_conversion_cast %[[LHS]] : complex<f32> to ![[C_TY:.*>]]
-// CHECK: %[[CASTED_RHS:.*]] = unrealized_conversion_cast %[[RHS]] : complex<f32> to ![[C_TY]]
+// CHECK: %[[CASTED_LHS:.*]] = builtin.unrealized_conversion_cast %[[LHS]] : complex<f32> to ![[C_TY:.*>]]
+// CHECK: %[[CASTED_RHS:.*]] = builtin.unrealized_conversion_cast %[[RHS]] : complex<f32> to ![[C_TY]]
 
 // CHECK: %[[LHS_RE:.*]] = llvm.extractvalue %[[CASTED_LHS]][0] : ![[C_TY]]
 // CHECK: %[[LHS_IM:.*]] = llvm.extractvalue %[[CASTED_LHS]][1] : ![[C_TY]]
 // CHECK: %[[RHS_RE:.*]] = llvm.extractvalue %[[CASTED_RHS]][0] : ![[C_TY]]
 // CHECK: %[[RHS_IM:.*]] = llvm.extractvalue %[[CASTED_RHS]][1] : ![[C_TY]]
 // CHECK: %[[RESULT_0:.*]] = llvm.mlir.undef : ![[C_TY]]
 
 // CHECK-DAG: %[[REAL_TMP_0:.*]] = llvm.fmul %[[RHS_RE]], %[[LHS_RE]]  : f32
 // CHECK-DAG: %[[REAL_TMP_1:.*]] = llvm.fmul %[[RHS_IM]], %[[LHS_IM]]  : f32
 // CHECK: %[[REAL:.*]] = llvm.fsub %[[REAL_TMP_0]], %[[REAL_TMP_1]]  : f32
 
 // CHECK-DAG: %[[IMAG_TMP_0:.*]] = llvm.fmul %[[LHS_IM]], %[[RHS_RE]]  : f32
 // CHECK-DAG: %[[IMAG_TMP_1:.*]] = llvm.fmul %[[LHS_RE]], %[[RHS_IM]]  : f32
 // CHECK: %[[IMAG:.*]] = llvm.fadd %[[IMAG_TMP_0]], %[[IMAG_TMP_1]]  : f32
 
 // CHECK: %[[RESULT_1:.*]] = llvm.insertvalue %[[REAL]], %[[RESULT_0]][0]
 // CHECK: %[[RESULT_2:.*]] = llvm.insertvalue %[[IMAG]], %[[RESULT_1]][1]
 
-// CHECK: %[[CASTED_RESULT:.*]] = unrealized_conversion_cast %[[RESULT_2]] : ![[C_TY]] to complex<f32>
+// CHECK: %[[CASTED_RESULT:.*]] = builtin.unrealized_conversion_cast %[[RESULT_2]] : ![[C_TY]] to complex<f32>
 // CHECK: return %[[CASTED_RESULT]] : complex<f32>
 
 // CHECK-LABEL: func @complex_abs
 // CHECK-SAME: %[[ARG:.*]]: complex<f32>
 func @complex_abs(%arg: complex<f32>) -> f32 {
   %abs = complex.abs %arg: complex<f32>
   return %abs : f32
 }
-// CHECK: %[[CASTED_ARG:.*]] = unrealized_conversion_cast %[[ARG]] : complex<f32> to ![[C_TY:.*>]]
+// CHECK: %[[CASTED_ARG:.*]] = builtin.unrealized_conversion_cast %[[ARG]] : complex<f32> to ![[C_TY:.*>]]
 // CHECK: %[[REAL:.*]] = llvm.extractvalue %[[CASTED_ARG]][0] : ![[C_TY]]
 // CHECK: %[[IMAG:.*]] = llvm.extractvalue %[[CASTED_ARG]][1] : ![[C_TY]]
 // CHECK-DAG: %[[REAL_SQ:.*]] = llvm.fmul %[[REAL]], %[[REAL]]  : f32
 // CHECK-DAG: %[[IMAG_SQ:.*]] = llvm.fmul %[[IMAG]], %[[IMAG]]  : f32
 // CHECK: %[[SQ_NORM:.*]] = llvm.fadd %[[REAL_SQ]], %[[IMAG_SQ]]  : f32
 // CHECK: %[[NORM:.*]] = "llvm.intr.sqrt"(%[[SQ_NORM]]) : (f32) -> f32
 // CHECK: return %[[NORM]] : f32
 

mlir/test/Conversion/MemRefToLLVM/convert-dynamic-memref-ops.mlir

 // RUN: mlir-opt -split-input-file -convert-memref-to-llvm %s | FileCheck %s
 // RUN: mlir-opt -split-input-file -convert-memref-to-llvm='use-aligned-alloc=1' %s | FileCheck %s --check-prefix=ALIGNED-ALLOC
 
 // CHECK-LABEL: func @mixed_alloc(
 //       CHECK:   %[[Marg:.*]]: index, %[[Narg:.*]]: index)
 func @mixed_alloc(%arg0: index, %arg1: index) -> memref<?x42x?xf32> {
-//       CHECK:  %[[M:.*]] = unrealized_conversion_cast %[[Marg]]
-//       CHECK:  %[[N:.*]] = unrealized_conversion_cast %[[Narg]]
+//       CHECK:  %[[M:.*]] = builtin.unrealized_conversion_cast %[[Marg]]
+//       CHECK:  %[[N:.*]] = builtin.unrealized_conversion_cast %[[Narg]]
 //       CHECK:  %[[c42:.*]] = llvm.mlir.constant(42 : index) : i64
 //  CHECK-NEXT:  %[[one:.*]] = llvm.mlir.constant(1 : index) : i64
 //  CHECK-NEXT:  %[[st0:.*]] = llvm.mul %[[N]], %[[c42]] : i64
 //  CHECK-NEXT:  %[[sz:.*]] = llvm.mul %[[st0]], %[[M]] : i64
 //  CHECK-NEXT:  %[[null:.*]] = llvm.mlir.null : !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[gep:.*]] = llvm.getelementptr %[[null]][%[[sz]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[sz_bytes:.*]] = llvm.ptrtoint %[[gep]] : !llvm.ptr<f32> to i64
 //  CHECK-NEXT:  llvm.call @malloc(%[[sz_bytes]]) : (i64) -> !llvm.ptr<i8>
// CHECK-NEXT:  llvm.call @free(%[[ptri8]]) : (!llvm.ptr<i8>) -> ()
   return
 }
 
 // -----
 
 // CHECK-LABEL: func @dynamic_alloc(
 //       CHECK:   %[[Marg:.*]]: index, %[[Narg:.*]]: index)
 func @dynamic_alloc(%arg0: index, %arg1: index) -> memref<?x?xf32> {
-//       CHECK:  %[[M:.*]] = unrealized_conversion_cast %[[Marg]]
-//       CHECK:  %[[N:.*]] = unrealized_conversion_cast %[[Narg]]
+//       CHECK:  %[[M:.*]] = builtin.unrealized_conversion_cast %[[Marg]]
+//       CHECK:  %[[N:.*]] = builtin.unrealized_conversion_cast %[[Narg]]
 //  CHECK-NEXT:  %[[one:.*]] = llvm.mlir.constant(1 : index) : i64
 //  CHECK-NEXT:  %[[sz:.*]] = llvm.mul %[[N]], %[[M]] : i64
 //  CHECK-NEXT:  %[[null:.*]] = llvm.mlir.null : !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[gep:.*]] = llvm.getelementptr %[[null]][%[[sz]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[sz_bytes:.*]] = llvm.ptrtoint %[[gep]] : !llvm.ptr<f32> to i64
 //  CHECK-NEXT:  llvm.call @malloc(%[[sz_bytes]]) : (i64) -> !llvm.ptr<i8>
 //  CHECK-NEXT:  llvm.bitcast %{{.*}} : !llvm.ptr<i8> to !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
//  CHECK-NEXT:  llvm.insertvalue %[[one]], %{{.*}}[4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   return %0 : memref<?x?xf32>
 }
 
 // -----
 
 // CHECK-LABEL: func @dynamic_alloca
 // CHECK: %[[Marg:.*]]: index, %[[Narg:.*]]: index)
 func @dynamic_alloca(%arg0: index, %arg1: index) -> memref<?x?xf32> {
-//       CHECK:  %[[M:.*]] = unrealized_conversion_cast %[[Marg]]
-//       CHECK:  %[[N:.*]] = unrealized_conversion_cast %[[Narg]]
+//       CHECK:  %[[M:.*]] = builtin.unrealized_conversion_cast %[[Marg]]
+//       CHECK:  %[[N:.*]] = builtin.unrealized_conversion_cast %[[Narg]]
 //  CHECK-NEXT:  %[[st1:.*]] = llvm.mlir.constant(1 : index) : i64
 //  CHECK-NEXT:  %[[num_elems:.*]] = llvm.mul %[[N]], %[[M]] : i64
 //  CHECK-NEXT:  %[[null:.*]] = llvm.mlir.null : !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[gep:.*]] = llvm.getelementptr %[[null]][%[[num_elems]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  %[[sz_bytes:.*]] = llvm.ptrtoint %[[gep]] : !llvm.ptr<f32> to i64
 //  CHECK-NEXT:  %[[allocated:.*]] = llvm.alloca %[[sz_bytes]] x f32 : (i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  llvm.insertvalue %[[allocated]], %{{.*}}[0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
// ALIGNED-ALLOC-NEXT:  llvm.bitcast %[[allocated]] : !llvm.ptr<i8> to !llvm.ptr<f32>
   return %0 : memref<32x18xf32>
 }
 
 // -----
 
 // CHECK-LABEL: func @mixed_load(
 // CHECK:         %{{.*}}, %[[Iarg:.*]]: index, %[[Jarg:.*]]: index)
 func @mixed_load(%mixed : memref<42x?xf32>, %i : index, %j : index) {
-//       CHECK:  %[[I:.*]] = unrealized_conversion_cast %[[Iarg]]
-//       CHECK:  %[[J:.*]] = unrealized_conversion_cast %[[Jarg]]
+//       CHECK:  %[[I:.*]] = builtin.unrealized_conversion_cast %[[Iarg]]
+//       CHECK:  %[[J:.*]] = builtin.unrealized_conversion_cast %[[Jarg]]
 //       CHECK:  %[[ptr:.*]] = llvm.extractvalue %[[ld:.*]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[st0:.*]] = llvm.extractvalue %[[ld]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[offI:.*]] = llvm.mul %[[I]], %[[st0]] : i64
 //  CHECK-NEXT:  %[[off1:.*]] = llvm.add %[[offI]], %[[J]] : i64
 //  CHECK-NEXT:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.load %[[addr]] : !llvm.ptr<f32>
   %0 = memref.load %mixed[%i, %j] : memref<42x?xf32>
   return
 }
 
 // -----
 
 // CHECK-LABEL: func @dynamic_load(
 // CHECK:         %{{.*}}, %[[Iarg:.*]]: index, %[[Jarg:.*]]: index)
 func @dynamic_load(%dynamic : memref<?x?xf32>, %i : index, %j : index) {
-//       CHECK:  %[[I:.*]] = unrealized_conversion_cast %[[Iarg]]
-//       CHECK:  %[[J:.*]] = unrealized_conversion_cast %[[Jarg]]
+//       CHECK:  %[[I:.*]] = builtin.unrealized_conversion_cast %[[Iarg]]
+//       CHECK:  %[[J:.*]] = builtin.unrealized_conversion_cast %[[Jarg]]
 //       CHECK:  %[[ptr:.*]] = llvm.extractvalue %[[ld:.*]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[st0:.*]] = llvm.extractvalue %[[ld]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[offI:.*]] = llvm.mul %[[I]], %[[st0]] : i64
 //  CHECK-NEXT:  %[[off1:.*]] = llvm.add %[[offI]], %[[J]] : i64
 //  CHECK-NEXT:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.load %[[addr]] : !llvm.ptr<f32>
   %0 = memref.load %dynamic[%i, %j] : memref<?x?xf32>
   return
 }
 
 // -----
 
 // CHECK-LABEL: func @prefetch
 // CHECK:         %{{.*}}, %[[Iarg:.*]]: index, %[[Jarg:.*]]: index)
 func @prefetch(%A : memref<?x?xf32>, %i : index, %j : index) {
-//      CHECK:  %[[I:.*]] = unrealized_conversion_cast %[[Iarg]]
-//      CHECK:  %[[J:.*]] = unrealized_conversion_cast %[[Jarg]]
+//      CHECK:  %[[I:.*]] = builtin.unrealized_conversion_cast %[[Iarg]]
+//      CHECK:  %[[J:.*]] = builtin.unrealized_conversion_cast %[[Jarg]]
 //      CHECK:  %[[ptr:.*]] = llvm.extractvalue %[[ld:.*]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 // CHECK-NEXT:  %[[st0:.*]] = llvm.extractvalue %[[ld]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 // CHECK-NEXT:  %[[offI:.*]] = llvm.mul %[[I]], %[[st0]] : i64
 // CHECK-NEXT:  %[[off1:.*]] = llvm.add %[[offI]], %[[J]] : i64
 // CHECK-NEXT:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK-NEXT:  [[C1:%.*]] = llvm.mlir.constant(1 : i32) : i32
 // CHECK-NEXT:  [[C3:%.*]] = llvm.mlir.constant(3 : i32) : i32
 // CHECK-NEXT:  [[C1_1:%.*]] = llvm.mlir.constant(1 : i32) : i32
// CHECK:  "llvm.intr.prefetch"(%{{.*}}, [[C0_2]], [[C2]], [[C0_3]]) : (!llvm.ptr<f32>, i32, i32, i32) -> ()
   return
 }
 
 // -----
 
 // CHECK-LABEL: func @dynamic_store
 // CHECK:         %{{.*}}, %[[Iarg:.*]]: index, %[[Jarg:.*]]: index
 func @dynamic_store(%dynamic : memref<?x?xf32>, %i : index, %j : index, %val : f32) {
-//       CHECK:  %[[I:.*]] = unrealized_conversion_cast %[[Iarg]]
-//       CHECK:  %[[J:.*]] = unrealized_conversion_cast %[[Jarg]]
+//       CHECK:  %[[I:.*]] = builtin.unrealized_conversion_cast %[[Iarg]]
+//       CHECK:  %[[J:.*]] = builtin.unrealized_conversion_cast %[[Jarg]]
 //       CHECK:  %[[ptr:.*]] = llvm.extractvalue %[[ld:.*]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[st0:.*]] = llvm.extractvalue %[[ld]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[offI:.*]] = llvm.mul %[[I]], %[[st0]] : i64
 //  CHECK-NEXT:  %[[off1:.*]] = llvm.add %[[offI]], %[[J]] : i64
 //  CHECK-NEXT:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.store %{{.*}}, %[[addr]] : !llvm.ptr<f32>
   memref.store %val, %dynamic[%i, %j] : memref<?x?xf32>
   return
 }
 
 // -----
 
 // CHECK-LABEL: func @mixed_store
 // CHECK:         %{{.*}}, %[[Iarg:.*]]: index, %[[Jarg:.*]]: index
 func @mixed_store(%mixed : memref<42x?xf32>, %i : index, %j : index, %val : f32) {
-//       CHECK:  %[[I:.*]] = unrealized_conversion_cast %[[Iarg]]
-//       CHECK:  %[[J:.*]] = unrealized_conversion_cast %[[Jarg]]
+//       CHECK:  %[[I:.*]] = builtin.unrealized_conversion_cast %[[Iarg]]
+//       CHECK:  %[[J:.*]] = builtin.unrealized_conversion_cast %[[Jarg]]
 //       CHECK:  %[[ptr:.*]] = llvm.extractvalue %[[ld:.*]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[st0:.*]] = llvm.extractvalue %[[ld]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 //  CHECK-NEXT:  %[[offI:.*]] = llvm.mul %[[I]], %[[st0]] : i64
 //  CHECK-NEXT:  %[[off1:.*]] = llvm.add %[[offI]], %[[J]] : i64
 //  CHECK-NEXT:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 //  CHECK-NEXT:  llvm.store %{{.*}}, %[[addr]] : !llvm.ptr<f32>
   memref.store %val, %mixed[%i, %j] : memref<42x?xf32>
   return

 // CHECK-LABEL: @memref_dim_with_dyn_index
 // CHECK: %{{.*}}, %[[IDXarg:.*]]: index
 func @memref_dim_with_dyn_index(%arg : memref<3x?xf32>, %idx : index) -> index {
   // CHECK-DAG: %[[C0:.*]] = llvm.mlir.constant(0 : index) : i64
   // CHECK-DAG: %[[C1:.*]] = llvm.mlir.constant(1 : index) : i64
   // CHECK-DAG: %[[SIZES:.*]] = llvm.extractvalue %{{.*}}[3] : ![[DESCR_TY:.*]]
   // CHECK-DAG: %[[SIZES_PTR:.*]] = llvm.alloca %[[C1]] x !llvm.array<2 x i64> : (i64) -> !llvm.ptr<array<2 x i64>>
   // CHECK-DAG: llvm.store %[[SIZES]], %[[SIZES_PTR]] : !llvm.ptr<array<2 x i64>>
-  // CHECK-DAG: %[[IDX:.*]] = unrealized_conversion_cast %[[IDXarg]]
+  // CHECK-DAG: %[[IDX:.*]] = builtin.unrealized_conversion_cast %[[IDXarg]]
   // CHECK-DAG: %[[RESULT_PTR:.*]] = llvm.getelementptr %[[SIZES_PTR]][%[[C0]], %[[IDX]]] : (!llvm.ptr<array<2 x i64>>, i64, i64) -> !llvm.ptr<i64>
   // CHECK-DAG: %[[RESULT:.*]] = llvm.load %[[RESULT_PTR]] : !llvm.ptr<i64>
   %result = memref.dim %arg, %idx : memref<3x?xf32>
   return %result : index
 }
 
 // -----
 
 // CHECK-LABEL: @memref_reinterpret_cast_ranked_to_static_shape
 func @memref_reinterpret_cast_ranked_to_static_shape(%input : memref<2x3xf32>) {
   %output = memref.reinterpret_cast %input to
            offset: [0], sizes: [6, 1], strides: [1, 1]
            : memref<2x3xf32> to memref<6x1xf32>
   return
 }
-// CHECK: [[INPUT:%.*]] = unrealized_conversion_cast %{{.*}} :
+// CHECK: [[INPUT:%.*]] = builtin.unrealized_conversion_cast %{{.*}} :
 // CHECK: to [[TY:!.*]]
 // CHECK: [[OUT_0:%.*]] = llvm.mlir.undef : [[TY]]
 // CHECK: [[BASE_PTR:%.*]] = llvm.extractvalue [[INPUT]][0] : [[TY]]
 // CHECK: [[ALIGNED_PTR:%.*]] = llvm.extractvalue [[INPUT]][1] : [[TY]]
 // CHECK: [[OUT_1:%.*]] = llvm.insertvalue [[BASE_PTR]], [[OUT_0]][0] : [[TY]]
 // CHECK: [[OUT_2:%.*]] = llvm.insertvalue [[ALIGNED_PTR]], [[OUT_1]][1] : [[TY]]
 // CHECK: [[OFFSET:%.*]] = llvm.mlir.constant(0 : index) : i64
 // CHECK: [[OUT_3:%.*]] = llvm.insertvalue [[OFFSET]], [[OUT_2]][2] : [[TY]]
%output = memref.reinterpret_cast %input to
            offset: [%offset], sizes: [%size_0, %size_1],
            strides: [%stride_0, %stride_1]
            : memref<*xf32> to memref<?x?xf32, offset: ?, strides: [?, ?]>
   return
 }
 // CHECK-SAME: ([[OFFSETarg:%[a-z,0-9]+]]: index,
 // CHECK-SAME: [[SIZE_0arg:%[a-z,0-9]+]]: index, [[SIZE_1arg:%[a-z,0-9]+]]: index,
 // CHECK-SAME: [[STRIDE_0arg:%[a-z,0-9]+]]: index, [[STRIDE_1arg:%[a-z,0-9]+]]: index,
-// CHECK: [[INPUT:%.*]] = unrealized_conversion_cast
-// CHECK: [[OFFSET:%.*]] = unrealized_conversion_cast [[OFFSETarg]]
-// CHECK: [[SIZE_0:%.*]] = unrealized_conversion_cast [[SIZE_0arg]]
-// CHECK: [[SIZE_1:%.*]] = unrealized_conversion_cast [[SIZE_1arg]]
-// CHECK: [[STRIDE_0:%.*]] = unrealized_conversion_cast [[STRIDE_0arg]]
-// CHECK: [[STRIDE_1:%.*]] = unrealized_conversion_cast [[STRIDE_1arg]]
+// CHECK: [[INPUT:%.*]] = builtin.unrealized_conversion_cast
+// CHECK: [[OFFSET:%.*]] = builtin.unrealized_conversion_cast [[OFFSETarg]]
+// CHECK: [[SIZE_0:%.*]] = builtin.unrealized_conversion_cast [[SIZE_0arg]]
+// CHECK: [[SIZE_1:%.*]] = builtin.unrealized_conversion_cast [[SIZE_1arg]]
+// CHECK: [[STRIDE_0:%.*]] = builtin.unrealized_conversion_cast [[STRIDE_0arg]]
+// CHECK: [[STRIDE_1:%.*]] = builtin.unrealized_conversion_cast [[STRIDE_1arg]]
 // CHECK: [[OUT_0:%.*]] = llvm.mlir.undef : [[TY:!.*]]
 // CHECK: [[DESCRIPTOR:%.*]] = llvm.extractvalue [[INPUT]][1] : !llvm.struct<(i64, ptr<i8>)>
 // CHECK: [[BASE_PTR_PTR:%.*]] = llvm.bitcast [[DESCRIPTOR]] : !llvm.ptr<i8> to !llvm.ptr<ptr<f32>>
 // CHECK: [[BASE_PTR:%.*]] = llvm.load [[BASE_PTR_PTR]] : !llvm.ptr<ptr<f32>>
 // CHECK: [[BASE_PTR_PTR_:%.*]] = llvm.bitcast [[DESCRIPTOR]] : !llvm.ptr<i8> to !llvm.ptr<ptr<f32>>
 // CHECK: [[C1:%.*]] = llvm.mlir.constant(1 : index) : i64
 // CHECK: [[ALIGNED_PTR_PTR:%.*]] = llvm.getelementptr [[BASE_PTR_PTR_]]{{\[}}[[C1]]]
 // CHECK-SAME: : (!llvm.ptr<ptr<f32>>, i64) -> !llvm.ptr<ptr<f32>>

 // -----
 
 // CHECK-LABEL: @memref_reshape
 func @memref_reshape(%input : memref<2x3xf32>, %shape : memref<?xindex>) {
   %output = memref.reshape %input(%shape)
                 : (memref<2x3xf32>, memref<?xindex>) -> memref<*xf32>
   return
 }
-// CHECK: [[INPUT:%.*]] = unrealized_conversion_cast %{{.*}} to [[INPUT_TY:!.*]]
-// CHECK: [[SHAPE:%.*]] = unrealized_conversion_cast %{{.*}} to [[SHAPE_TY:!.*]]
+// CHECK: [[INPUT:%.*]] = builtin.unrealized_conversion_cast %{{.*}} to [[INPUT_TY:!.*]]
+// CHECK: [[SHAPE:%.*]] = builtin.unrealized_conversion_cast %{{.*}} to [[SHAPE_TY:!.*]]
 // CHECK: [[RANK:%.*]] = llvm.extractvalue [[SHAPE]][3, 0] : [[SHAPE_TY]]
 // CHECK: [[UNRANKED_OUT_O:%.*]] = llvm.mlir.undef : !llvm.struct<(i64, ptr<i8>)>
 // CHECK: [[UNRANKED_OUT_1:%.*]] = llvm.insertvalue [[RANK]], [[UNRANKED_OUT_O]][0] : !llvm.struct<(i64, ptr<i8>)>
 
 // Compute size in bytes to allocate result ranked descriptor
 // CHECK: [[C1:%.*]] = llvm.mlir.constant(1 : index) : i64
 // CHECK: [[C2:%.*]] = llvm.mlir.constant(2 : index) : i64
 // CHECK: [[PTR_SIZE:%.*]] = llvm.mlir.constant(8 : index) : i64

mlir/test/Conversion/MemRefToLLVM/convert-static-memref-ops.mlir

 
 // -----
 
 // CHECK-LABEL: func @static_load
 // CHECK:         %[[MEMREF:.*]]: memref<10x42xf32>,
 // CHECK:         %[[I:.*]]: index,
 // CHECK:         %[[J:.*]]: index)
 func @static_load(%static : memref<10x42xf32>, %i : index, %j : index) {
-// CHECK:  %[[II:.*]] = unrealized_conversion_cast %[[I]]
-// CHECK:  %[[JJ:.*]] = unrealized_conversion_cast %[[J]]
+// CHECK:  %[[II:.*]] = builtin.unrealized_conversion_cast %[[I]]
+// CHECK:  %[[JJ:.*]] = builtin.unrealized_conversion_cast %[[J]]
 // CHECK:  %[[ptr:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 // CHECK:  %[[st0:.*]] = llvm.mlir.constant(42 : index) : i64
 // CHECK:  %[[offI:.*]] = llvm.mul %[[II]], %[[st0]] : i64
 // CHECK:  %[[off1:.*]] = llvm.add %[[offI]], %[[JJ]] : i64
 // CHECK:  %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK:  llvm.load %[[addr]] : !llvm.ptr<f32>
   %0 = memref.load %static[%i, %j] : memref<10x42xf32>
   return

 }
 
 // -----
 
 // CHECK-LABEL: func @static_store
 // CHECK:         %[[MEMREF:.*]]: memref<10x42xf32>,
 // CHECK-SAME:    %[[I:.*]]: index, %[[J:.*]]: index,
 func @static_store(%static : memref<10x42xf32>, %i : index, %j : index, %val : f32) {
-// CHECK: %[[II:.*]] = unrealized_conversion_cast %[[I]]
-// CHECK: %[[JJ:.*]] = unrealized_conversion_cast %[[J]]
+// CHECK: %[[II:.*]] = builtin.unrealized_conversion_cast %[[I]]
+// CHECK: %[[JJ:.*]] = builtin.unrealized_conversion_cast %[[J]]
 // CHECK: %[[ptr:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
 // CHECK: %[[st0:.*]] = llvm.mlir.constant(42 : index) : i64
 // CHECK: %[[offI:.*]] = llvm.mul %[[II]], %[[st0]] : i64
 // CHECK: %[[off1:.*]] = llvm.add %[[offI]], %[[JJ]] : i64
 // CHECK: %[[addr:.*]] = llvm.getelementptr %[[ptr]][%[[off1]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK: llvm.store %{{.*}}, %[[addr]] : !llvm.ptr<f32>
 
   memref.store %val, %static[%i, %j] : memref<10x42xf32>

 
 // Check that consistent types are emitted in address arithemic in presence of
 // a data layout specification.
 module attributes { dlti.dl_spec = #dlti.dl_spec<#dlti.dl_entry<index, 32>> } {
   func @address() {
     %c1 = constant 1 : index
     %0 = memref.alloc(%c1) : memref<? x vector<2xf32>>
     // CHECK: %[[CST_S:.*]] = constant 1 : index
-    // CHECK: %[[CST:.*]] = unrealized_conversion_cast
+    // CHECK: %[[CST:.*]] = builtin.unrealized_conversion_cast
     // CHECK: llvm.mlir.null
     // CHECK: llvm.getelementptr %{{.*}}[[CST]]
     // CHECK: llvm.ptrtoint %{{.*}} : !llvm.ptr<{{.*}}> to i32
     // CHECK: llvm.ptrtoint %{{.*}} : !llvm.ptr<{{.*}}> to i32
     // CHECK: llvm.add %{{.*}} : i32
     // CHECK: llvm.call @malloc(%{{.*}}) : (i32) -> !llvm.ptr
     // CHECK: llvm.ptrtoint %{{.*}} : !llvm.ptr<{{.*}}> to i32
     // CHECK: llvm.sub {{.*}} : i32
     // CHECK: llvm.add {{.*}} : i32
     // CHECK: llvm.urem {{.*}} : i32
     // CHECK: llvm.sub {{.*}} : i32
     // CHECK: llvm.inttoptr %{{.*}} : i32 to !llvm.ptr
     return
   }
 }
 

mlir/test/Conversion/MemRefToLLVM/memref-to-llvm.mlir

 // RUN: mlir-opt -convert-memref-to-llvm %s -split-input-file | FileCheck %s
 // RUN: mlir-opt -convert-memref-to-llvm='index-bitwidth=32' %s -split-input-file | FileCheck --check-prefix=CHECK32 %s
 
 
 // CHECK-LABEL: func @view(
 // CHECK: %[[ARG0F:.*]]: index, %[[ARG1F:.*]]: index, %[[ARG2F:.*]]: index
 func @view(%arg0 : index, %arg1 : index, %arg2 : index) {
   // CHECK: llvm.mlir.constant(2048 : index) : i64
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<i8>, ptr<i8>, i64, array<1 x i64>, array<1 x i64>)>
   %0 = memref.alloc() : memref<2048xi8>
 
   // Test two dynamic sizes.
-  // CHECK: %[[ARG2:.*]] = unrealized_conversion_cast %[[ARG2F:.*]]
-  // CHECK: %[[ARG0:.*]] = unrealized_conversion_cast %[[ARG0F:.*]]
-  // CHECK: %[[ARG1:.*]] = unrealized_conversion_cast %[[ARG1F:.*]]
+  // CHECK: %[[ARG2:.*]] = builtin.unrealized_conversion_cast %[[ARG2F:.*]]
+  // CHECK: %[[ARG0:.*]] = builtin.unrealized_conversion_cast %[[ARG0F:.*]]
+  // CHECK: %[[ARG1:.*]] = builtin.unrealized_conversion_cast %[[ARG1F:.*]]
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BASE_PTR:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<i8>, ptr<i8>, i64, array<1 x i64>, array<1 x i64>)>
   // CHECK: %[[SHIFTED_BASE_PTR:.*]] = llvm.getelementptr %[[BASE_PTR]][%[[ARG2]]] : (!llvm.ptr<i8>, i64) -> !llvm.ptr<i8>
   // CHECK: %[[CAST_SHIFTED_BASE_PTR:.*]] = llvm.bitcast %[[SHIFTED_BASE_PTR]] : !llvm.ptr<i8> to !llvm.ptr<f32>
   // CHECK: llvm.insertvalue %[[CAST_SHIFTED_BASE_PTR]], %{{.*}}[1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[C0:.*]] = llvm.mlir.constant(0 : index) : i64
   // CHECK: llvm.insertvalue %[[C0]], %{{.*}}[2] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.insertvalue %[[ARG1]], %{{.*}}[3, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(1 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.insertvalue %[[ARG0]], %{{.*}}[3, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mul %{{.*}}, %[[ARG1]]
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   %1 = memref.view %0[%arg2][%arg0, %arg1] : memref<2048xi8> to memref<?x?xf32>
 
   // Test one dynamic size.
-  // CHECK: %[[ARG2:.*]] = unrealized_conversion_cast %[[ARG2F:.*]]
-  // CHECK: %[[ARG1:.*]] = unrealized_conversion_cast %[[ARG1F:.*]]
+  // CHECK: %[[ARG2:.*]] = builtin.unrealized_conversion_cast %[[ARG2F:.*]]
+  // CHECK: %[[ARG1:.*]] = builtin.unrealized_conversion_cast %[[ARG1F:.*]]
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BASE_PTR_2:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<i8>, ptr<i8>, i64, array<1 x i64>, array<1 x i64>)>
   // CHECK: %[[SHIFTED_BASE_PTR_2:.*]] = llvm.getelementptr %[[BASE_PTR_2]][%[[ARG2]]] : (!llvm.ptr<i8>, i64) -> !llvm.ptr<i8>
   // CHECK: %[[CAST_SHIFTED_BASE_PTR_2:.*]] = llvm.bitcast %[[SHIFTED_BASE_PTR_2]] : !llvm.ptr<i8> to !llvm.ptr<f32>
   // CHECK: llvm.insertvalue %[[CAST_SHIFTED_BASE_PTR_2]], %{{.*}}[1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[C0_2:.*]] = llvm.mlir.constant(0 : index) : i64
   // CHECK: llvm.insertvalue %[[C0_2]], %{{.*}}[2] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.insertvalue %[[ARG1]], %{{.*}}[3, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(1 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(4 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[3, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mul %{{.*}}, %[[ARG1]]
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   %3 = memref.view %0[%arg2][%arg1] : memref<2048xi8> to memref<4x?xf32>
 
   // Test static sizes.
-  // CHECK: %[[ARG2:.*]] = unrealized_conversion_cast %[[ARG2F:.*]]
+  // CHECK: %[[ARG2:.*]] = builtin.unrealized_conversion_cast %[[ARG2F:.*]]
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BASE_PTR_3:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<i8>, ptr<i8>, i64, array<1 x i64>, array<1 x i64>)>
   // CHECK: %[[SHIFTED_BASE_PTR_3:.*]] = llvm.getelementptr %[[BASE_PTR_3]][%[[ARG2]]] : (!llvm.ptr<i8>, i64) -> !llvm.ptr<i8>
   // CHECK: %[[CAST_SHIFTED_BASE_PTR_3:.*]] = llvm.bitcast %[[SHIFTED_BASE_PTR_3]] : !llvm.ptr<i8> to !llvm.ptr<f32>
   // CHECK: llvm.insertvalue %[[CAST_SHIFTED_BASE_PTR_3]], %{{.*}}[1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[C0_3:.*]] = llvm.mlir.constant(0 : index) : i64
   // CHECK: llvm.insertvalue %[[C0_3]], %{{.*}}[2] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(4 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[3, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(1 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(64 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[3, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(4 : index) : i64
   // CHECK: llvm.insertvalue %{{.*}}, %{{.*}}[4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   %5 = memref.view %0[%arg2][] : memref<2048xi8> to memref<64x4xf32>
 
   // Test view memory space.
   // CHECK: llvm.mlir.constant(2048 : index) : i64
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<i8, 4>, ptr<i8, 4>, i64, array<1 x i64>, array<1 x i64>)>
   %6 = memref.alloc() : memref<2048xi8, 4>
 
-  // CHECK: %[[ARG2:.*]] = unrealized_conversion_cast %[[ARG2F:.*]]
+  // CHECK: %[[ARG2:.*]] = builtin.unrealized_conversion_cast %[[ARG2F:.*]]
   // CHECK: llvm.mlir.undef : !llvm.struct<(ptr<f32, 4>, ptr<f32, 4>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BASE_PTR_4:.*]] = llvm.extractvalue %{{.*}}[1] : !llvm.struct<(ptr<i8, 4>, ptr<i8, 4>, i64, array<1 x i64>, array<1 x i64>)>
   // CHECK: %[[SHIFTED_BASE_PTR_4:.*]] = llvm.getelementptr %[[BASE_PTR_4]][%[[ARG2]]] : (!llvm.ptr<i8, 4>, i64) -> !llvm.ptr<i8, 4>
   // CHECK: %[[CAST_SHIFTED_BASE_PTR_4:.*]] = llvm.bitcast %[[SHIFTED_BASE_PTR_4]] : !llvm.ptr<i8, 4> to !llvm.ptr<f32, 4>
   // CHECK: llvm.insertvalue %[[CAST_SHIFTED_BASE_PTR_4]], %{{.*}}[1] : !llvm.struct<(ptr<f32, 4>, ptr<f32, 4>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[C0_4:.*]] = llvm.mlir.constant(0 : index) : i64
   // CHECK: llvm.insertvalue %[[C0_4]], %{{.*}}[2] : !llvm.struct<(ptr<f32, 4>, ptr<f32, 4>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: llvm.mlir.constant(4 : index) : i64

 // CHECK:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK:         %[[ARG2f:.*]]: index)
 // CHECK32-LABEL: func @subview(
 // CHECK32:         %[[MEM:.*]]: memref<{{.*}}>,
 // CHECK32:         %[[ARG0f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG2f:.*]]: index)
 func @subview(%0 : memref<64x4xf32, offset: 0, strides: [4, 1]>, %arg0 : index, %arg1 : index, %arg2 : index) {
-  // CHECK: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
+  // CHECK: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
 
-  // CHECK: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0c:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1c:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0c:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1c:.*]] = unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0c:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1c:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0c:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1c:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
 
   // CHECK: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>

 // CHECK:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK:         %[[ARG2f:.*]]: index)
 // CHECK32-LABEL: func @subview_non_zero_addrspace(
 // CHECK32:         %[[MEM:.*]]: memref<{{.*}}>,
 // CHECK32:         %[[ARG0f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG2f:.*]]: index)
 func @subview_non_zero_addrspace(%0 : memref<64x4xf32, offset: 0, strides: [4, 1], 3>, %arg0 : index, %arg1 : index, %arg2 : index) {
-  // CHECK: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
+  // CHECK: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
 
-  // CHECK: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0c:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1c:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0c:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1c:.*]] = unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0c:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1c:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0c:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1c:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
 
   // CHECK: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32, 3>, ptr<f32, 3>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32, 3> to !llvm.ptr<f32, 3>
   // CHECK: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32, 3>, ptr<f32, 3>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32, 3> to !llvm.ptr<f32, 3>
   // CHECK: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32, 3>, ptr<f32, 3>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32, 3>, ptr<f32, 3>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32, 3>, ptr<f32, 3>, i64, array<2 x i64>, array<2 x i64>)>

 // CHECK-SAME:         %[[ARG1f:[a-zA-Z0-9]*]]: index
 // CHECK-SAME:         %[[ARG2f:[a-zA-Z0-9]*]]: index
 // CHECK32-LABEL: func @subview_const_size(
 // CHECK32-SAME:         %[[MEM:.*]]: memref<{{.*}}>,
 // CHECK32-SAME:         %[[ARG0f:[a-zA-Z0-9]*]]: index
 // CHECK32-SAME:         %[[ARG1f:[a-zA-Z0-9]*]]: index
 // CHECK32-SAME:         %[[ARG2f:[a-zA-Z0-9]*]]: index
 func @subview_const_size(%0 : memref<64x4xf32, offset: 0, strides: [4, 1]>, %arg0 : index, %arg1 : index, %arg2 : index) {
-  // CHECK: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
+  // CHECK: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
 
-  // CHECK: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
 
   // CHECK: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>

 // CHECK-SAME:         %[[ARG1f:[a-zA-Z0-9]*]]: index
 // CHECK-SAME:         %[[ARG2f:[a-zA-Z0-9]*]]: index
 // CHECK32-LABEL: func @subview_const_stride(
 // CHECK32-SAME:         %[[MEM:.*]]: memref<{{.*}}>,
 // CHECK32-SAME:         %[[ARG0f:[a-zA-Z0-9]*]]: index
 // CHECK32-SAME:         %[[ARG1f:[a-zA-Z0-9]*]]: index
 // CHECK32-SAME:         %[[ARG2f:[a-zA-Z0-9]*]]: index
 func @subview_const_stride(%0 : memref<64x4xf32, offset: 0, strides: [4, 1]>, %arg0 : index, %arg1 : index, %arg2 : index) {
-  // CHECK: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
+  // CHECK: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
 
-  // CHECK: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0a:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1a:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG0b:.*]] = unrealized_conversion_cast %[[ARG0f]]
-  // CHECK32: %[[ARG1b:.*]] = unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0a:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1a:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG0b:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[ARG1b:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
 
   // CHECK: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>
   // CHECK: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i64, array<2 x i64>, array<2 x i64>)>

 }
 
 // -----
 
 // CHECK-LABEL: func @subview_const_stride_and_offset(
 // CHECK32-LABEL: func @subview_const_stride_and_offset(
 func @subview_const_stride_and_offset(%0 : memref<64x4xf32, offset: 0, strides: [4, 1]>) {
   // The last "insertvalue" that populates the memref descriptor from the function arguments.
-  // CHECK: %[[MEMREF:.*]] = unrealized_conversion_cast
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast
+  // CHECK: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast
 
   // CHECK32: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK32: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK32: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>

 // CHECK:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK:         %[[ARG2f:.*]]: index)
 // CHECK32-LABEL: func @subview_mixed_static_dynamic(
 // CHECK32:         %[[MEM:.*]]: memref<{{.*}}>,
 // CHECK32:         %[[ARG0f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG1f:[a-zA-Z0-9]*]]: index,
 // CHECK32:         %[[ARG2f:.*]]: index)
 func @subview_mixed_static_dynamic(%0 : memref<64x4xf32, offset: 0, strides: [4, 1]>, %arg0 : index, %arg1 : index, %arg2 : index) {
-  // CHECK32: %[[MEMREF:.*]] = unrealized_conversion_cast %[[MEM]]
-  // CHECK32: %[[ARG1:.*]] = unrealized_conversion_cast %[[ARG1f]]
-  // CHECK32: %[[ARG2:.*]] = unrealized_conversion_cast %[[ARG2f]]
-  // CHECK32: %[[ARG0:.*]] = unrealized_conversion_cast %[[ARG0f]]
+  // CHECK32: %[[MEMREF:.*]] = builtin.unrealized_conversion_cast %[[MEM]]
+  // CHECK32: %[[ARG1:.*]] = builtin.unrealized_conversion_cast %[[ARG1f]]
+  // CHECK32: %[[ARG2:.*]] = builtin.unrealized_conversion_cast %[[ARG2f]]
+  // CHECK32: %[[ARG0:.*]] = builtin.unrealized_conversion_cast %[[ARG0f]]
 
   // CHECK32: %[[DESC:.*]] = llvm.mlir.undef : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[BITCAST0:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK32: %[[DESC0:.*]] = llvm.insertvalue %[[BITCAST0]], %[[DESC]][0] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[BITCAST1:.*]] = llvm.bitcast %{{.*}} : !llvm.ptr<f32> to !llvm.ptr<f32>
   // CHECK32: %[[DESC1:.*]] = llvm.insertvalue %[[BITCAST1]], %[[DESC0]][1] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[STRIDE0:.*]] = llvm.extractvalue %[[MEMREF]][4, 0] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>
   // CHECK32: %[[STRIDE1:.*]] = llvm.extractvalue %[[MEMREF]][4, 1] : !llvm.struct<(ptr<f32>, ptr<f32>, i32, array<2 x i32>, array<2 x i32>)>

 
 // CHECK-LABEL: func @dim_of_unranked
 // CHECK32-LABEL: func @dim_of_unranked
 func @dim_of_unranked(%unranked: memref<*xi32>) -> index {
   %c0 = constant 0 : index
   %dim = memref.dim %unranked, %c0 : memref<*xi32>
   return %dim : index
 }
-// CHECK: %[[UNRANKED_DESC:.*]] = unrealized_conversion_cast
+// CHECK: %[[UNRANKED_DESC:.*]] = builtin.unrealized_conversion_cast
 
 // CHECK: %[[RANKED_DESC:.*]] = llvm.extractvalue %[[UNRANKED_DESC]][1]
 // CHECK-SAME:   : !llvm.struct<(i64, ptr<i8>)>
 
 // CHECK: %[[ZERO_D_DESC:.*]] = llvm.bitcast %[[RANKED_DESC]]
 // CHECK-SAME:   : !llvm.ptr<i8> to !llvm.ptr<struct<(ptr<i32>, ptr<i32>, i64)>>
 
 // CHECK: %[[C2_i32:.*]] = llvm.mlir.constant(2 : i32) : i32

mlir/test/Conversion/VectorToLLVM/vector-to-llvm.mlir

 
 func @bitcast_index_to_i8_vector(%input: vector<16xindex>) -> vector<128xi8> {
   %0 = vector.bitcast %input : vector<16xindex> to vector<128xi8>
   return %0 : vector<128xi8>
 }
 
 // CHECK-LABEL: @bitcast_index_to_i8_vector
 // CHECK-SAME:  %[[input:.*]]: vector<16xindex>
-// CHECK:       %[[T0:.*]] = unrealized_conversion_cast %[[input]] : vector<16xindex> to vector<16xi64>
+// CHECK:       %[[T0:.*]] = builtin.unrealized_conversion_cast %[[input]] : vector<16xindex> to vector<16xi64>
 // CHECK:       llvm.bitcast %[[T0]] : vector<16xi64> to vector<128xi8>
 
 // -----
 
 func @broadcast_vec1d_from_f32(%arg0: f32) -> vector<2xf32> {
   %0 = vector.broadcast %arg0 : f32 to vector<2xf32>
   return %0 : vector<2xf32>
 }

 
 func @broadcast_vec2d_from_vec1d(%arg0: vector<2xf32>) -> vector<3x2xf32> {
   %0 = vector.broadcast %arg0 : vector<2xf32> to vector<3x2xf32>
   return %0 : vector<3x2xf32>
 }
 // CHECK-LABEL: @broadcast_vec2d_from_vec1d(
 // CHECK-SAME:  %[[A:.*]]: vector<2xf32>)
 // CHECK:       %[[T0:.*]] = constant dense<0.000000e+00> : vector<3x2xf32>
-// CHECK:       %[[T1:.*]] = unrealized_conversion_cast %[[T0]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T1:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T2:.*]] = llvm.insertvalue %[[A]], %[[T1]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T3:.*]] = llvm.insertvalue %[[A]], %[[T2]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T4:.*]] = llvm.insertvalue %[[A]], %[[T3]][2] : !llvm.array<3 x vector<2xf32>>
-// CHECK:       %[[T5:.*]] = unrealized_conversion_cast %[[T4]] : !llvm.array<3 x vector<2xf32>> to vector<3x2xf32>
+// CHECK:       %[[T5:.*]] = builtin.unrealized_conversion_cast %[[T4]] : !llvm.array<3 x vector<2xf32>> to vector<3x2xf32>
 // CHECK:       return %[[T5]] : vector<3x2xf32>
 
 // -----
 
 func @broadcast_vec2d_from_index_vec1d(%arg0: vector<2xindex>) -> vector<3x2xindex> {
   %0 = vector.broadcast %arg0 : vector<2xindex> to vector<3x2xindex>
   return %0 : vector<3x2xindex>
 }
 // CHECK-LABEL: @broadcast_vec2d_from_index_vec1d(
 // CHECK-SAME:  %[[A:.*]]: vector<2xindex>)
 // CHECK:       %[[T0:.*]] = constant dense<0> : vector<3x2xindex>
-// CHECK:       %[[T1:.*]] = unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[T0]] : vector<3x2xindex> to !llvm.array<3 x vector<2xi64>>
+// CHECK:       %[[T1:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<3x2xindex> to !llvm.array<3 x vector<2xi64>>
 // CHECK:       %[[T3:.*]] = llvm.insertvalue %[[T1]], %[[T2]][0] : !llvm.array<3 x vector<2xi64>>
 
-// CHECK:       %[[T4:.*]] = unrealized_conversion_cast %{{.*}} : !llvm.array<3 x vector<2xi64>> to vector<3x2xindex>
+// CHECK:       %[[T4:.*]] = builtin.unrealized_conversion_cast %{{.*}} : !llvm.array<3 x vector<2xi64>> to vector<3x2xindex>
 // CHECK:       return %[[T4]] : vector<3x2xindex>
 
 // -----
 
 func @broadcast_vec3d_from_vec1d(%arg0: vector<2xf32>) -> vector<4x3x2xf32> {
   %0 = vector.broadcast %arg0 : vector<2xf32> to vector<4x3x2xf32>
   return %0 : vector<4x3x2xf32>
 }
 // CHECK-LABEL: @broadcast_vec3d_from_vec1d(
 // CHECK-SAME:  %[[A:.*]]: vector<2xf32>)
 // CHECK:       %[[T0:.*]] = constant dense<0.000000e+00> : vector<3x2xf32>
 // CHECK:       %[[T1:.*]] = constant dense<0.000000e+00> : vector<4x3x2xf32>
 
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[T0]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T3:.*]] = llvm.insertvalue %[[A]], %[[T2]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T4:.*]] = llvm.insertvalue %[[A]], %[[T3]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T5:.*]] = llvm.insertvalue %[[A]], %[[T4]][2] : !llvm.array<3 x vector<2xf32>>
 
-// CHECK:       %[[T6:.*]] = unrealized_conversion_cast %[[T1]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
+// CHECK:       %[[T6:.*]] = builtin.unrealized_conversion_cast %[[T1]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T7:.*]] = llvm.insertvalue %[[T5]], %[[T6]][0] : !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T8:.*]] = llvm.insertvalue %[[T5]], %[[T7]][1] : !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T9:.*]] = llvm.insertvalue %[[T5]], %[[T8]][2] : !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T10:.*]] = llvm.insertvalue %[[T5]], %[[T9]][3] : !llvm.array<4 x array<3 x vector<2xf32>>>
 
-// CHECK:       %[[T11:.*]] = unrealized_conversion_cast %[[T10]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
+// CHECK:       %[[T11:.*]] = builtin.unrealized_conversion_cast %[[T10]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
 // CHECK:       return %[[T11]] : vector<4x3x2xf32>
 
 // -----
 
 func @broadcast_vec3d_from_vec2d(%arg0: vector<3x2xf32>) -> vector<4x3x2xf32> {
   %0 = vector.broadcast %arg0 : vector<3x2xf32> to vector<4x3x2xf32>
   return %0 : vector<4x3x2xf32>
 }
 // CHECK-LABEL: @broadcast_vec3d_from_vec2d(
 // CHECK-SAME:  %[[A:.*]]: vector<3x2xf32>)
 // CHECK:       %[[T0:.*]] = constant dense<0.000000e+00> : vector<4x3x2xf32>
-// CHECK:       %[[T1:.*]] = unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[T0]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
+// CHECK:       %[[T1:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T3:.*]] = llvm.insertvalue %[[T1]], %[[T2]][0] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T4:.*]] = unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T4:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T5:.*]] = llvm.insertvalue %[[T4]], %[[T3]][1] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T6:.*]] = unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T6:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T7:.*]] = llvm.insertvalue %[[T6]], %[[T5]][2] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T8:.*]] = unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T8:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T9:.*]] = llvm.insertvalue %[[T8]], %[[T7]][3] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T10:.*]] = unrealized_conversion_cast %[[T9]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
+// CHECK:       %[[T10:.*]] = builtin.unrealized_conversion_cast %[[T9]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
 // CHECK:       return %[[T10]] : vector<4x3x2xf32>
 
 
 // -----
 
 func @broadcast_stretch(%arg0: vector<1xf32>) -> vector<4xf32> {
   %0 = vector.broadcast %arg0 : vector<1xf32> to vector<4xf32>
   return %0 : vector<4xf32>

 
 func @broadcast_stretch_at_start(%arg0: vector<1x4xf32>) -> vector<3x4xf32> {
   %0 = vector.broadcast %arg0 : vector<1x4xf32> to vector<3x4xf32>
   return %0 : vector<3x4xf32>
 }
 // CHECK-LABEL: @broadcast_stretch_at_start(
 // CHECK-SAME:  %[[A:.*]]: vector<1x4xf32>)
 // CHECK:       %[[T1:.*]] = constant dense<0.000000e+00> : vector<3x4xf32>
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[A]] : vector<1x4xf32> to !llvm.array<1 x vector<4xf32>>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<1x4xf32> to !llvm.array<1 x vector<4xf32>>
 // CHECK:       %[[T3:.*]] = llvm.extractvalue %[[T2]][0] : !llvm.array<1 x vector<4xf32>>
-// CHECK:       %[[T4:.*]] = unrealized_conversion_cast %[[T1]] : vector<3x4xf32> to !llvm.array<3 x vector<4xf32>>
+// CHECK:       %[[T4:.*]] = builtin.unrealized_conversion_cast %[[T1]] : vector<3x4xf32> to !llvm.array<3 x vector<4xf32>>
 // CHECK:       %[[T5:.*]] = llvm.insertvalue %[[T3]], %[[T4]][0] : !llvm.array<3 x vector<4xf32>>
 // CHECK:       %[[T6:.*]] = llvm.insertvalue %[[T3]], %[[T5]][1] : !llvm.array<3 x vector<4xf32>>
 // CHECK:       %[[T7:.*]] = llvm.insertvalue %[[T3]], %[[T6]][2] : !llvm.array<3 x vector<4xf32>>
-// CHECK:       %[[T8:.*]] = unrealized_conversion_cast %[[T7]] : !llvm.array<3 x vector<4xf32>> to vector<3x4xf32>
+// CHECK:       %[[T8:.*]] = builtin.unrealized_conversion_cast %[[T7]] : !llvm.array<3 x vector<4xf32>> to vector<3x4xf32>
 // CHECK:       return %[[T8]] : vector<3x4xf32>
 
 // -----
 
 func @broadcast_stretch_at_end(%arg0: vector<4x1xf32>) -> vector<4x3xf32> {
   %0 = vector.broadcast %arg0 : vector<4x1xf32> to vector<4x3xf32>
   return %0 : vector<4x3xf32>
 }
 // CHECK-LABEL: @broadcast_stretch_at_end(
 // CHECK-SAME:  %[[A:.*]]: vector<4x1xf32>)
 // CHECK:       %[[T1:.*]] = constant dense<0.000000e+00> : vector<4x3xf32>
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T3:.*]] = llvm.extractvalue %[[T2]][0] : !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T4:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T5:.*]] = llvm.extractelement %[[T3]]{{\[}}%[[T4]] : i64] : vector<1xf32>
 // CHECK:       %[[T6:.*]] = splat %[[T5]] : vector<3xf32>
-// CHECK:       %[[T7:.*]] = unrealized_conversion_cast %[[T1]] : vector<4x3xf32> to !llvm.array<4 x vector<3xf32>>
+// CHECK:       %[[T7:.*]] = builtin.unrealized_conversion_cast %[[T1]] : vector<4x3xf32> to !llvm.array<4 x vector<3xf32>>
 // CHECK:       %[[T8:.*]] = llvm.insertvalue %[[T6]], %[[T7]][0] : !llvm.array<4 x vector<3xf32>>
-// CHECK:       %[[T9:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
+// CHECK:       %[[T9:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T10:.*]] = llvm.extractvalue %[[T9]][1] : !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T11:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T12:.*]] = llvm.extractelement %[[T10]]{{\[}}%[[T11]] : i64] : vector<1xf32>
 // CHECK:       %[[T13:.*]] = splat %[[T12]] : vector<3xf32>
 // CHECK:       %[[T14:.*]] = llvm.insertvalue %[[T13]], %[[T8]][1] : !llvm.array<4 x vector<3xf32>>
-// CHECK:       %[[T15:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
+// CHECK:       %[[T15:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T16:.*]] = llvm.extractvalue %[[T15]][2] : !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T17:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T18:.*]] = llvm.extractelement %[[T16]]{{\[}}%[[T17]] : i64] : vector<1xf32>
 // CHECK:       %[[T19:.*]] = splat %[[T18]] : vector<3xf32>
 // CHECK:       %[[T20:.*]] = llvm.insertvalue %[[T19]], %[[T14]][2] : !llvm.array<4 x vector<3xf32>>
-// CHECK:       %[[T21:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
+// CHECK:       %[[T21:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1xf32> to !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T22:.*]] = llvm.extractvalue %[[T21]][3] : !llvm.array<4 x vector<1xf32>>
 // CHECK:       %[[T23:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T24:.*]] = llvm.extractelement %[[T22]]{{\[}}%[[T23]] : i64] : vector<1xf32>
 // CHECK:       %[[T25:.*]] = splat %[[T24]] : vector<3xf32>
 // CHECK:       %[[T26:.*]] = llvm.insertvalue %[[T25]], %[[T20]][3] : !llvm.array<4 x vector<3xf32>>
-// CHECK:       %[[T27:.*]] = unrealized_conversion_cast %[[T26]] : !llvm.array<4 x vector<3xf32>> to vector<4x3xf32>
+// CHECK:       %[[T27:.*]] = builtin.unrealized_conversion_cast %[[T26]] : !llvm.array<4 x vector<3xf32>> to vector<4x3xf32>
 // CHECK:       return %[[T27]] : vector<4x3xf32>
 
 // -----
 
 func @broadcast_stretch_in_middle(%arg0: vector<4x1x2xf32>) -> vector<4x3x2xf32> {
   %0 = vector.broadcast %arg0 : vector<4x1x2xf32> to vector<4x3x2xf32>
   return %0 : vector<4x3x2xf32>
 }
 // CHECK-LABEL: @broadcast_stretch_in_middle(
 // CHECK-SAME:  %[[A:.*]]: vector<4x1x2xf32>) -> vector<4x3x2xf32> {
 // CHECK:       %[[T1:.*]] = constant dense<0.000000e+00> : vector<4x3x2xf32>
 // CHECK:       %[[T2:.*]] = constant dense<0.000000e+00> : vector<3x2xf32>
-// CHECK:       %[[T3:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
+// CHECK:       %[[T3:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
 // CHECK:       %[[T4:.*]] = llvm.extractvalue %[[T3]][0, 0] : !llvm.array<4 x array<1 x vector<2xf32>>>
-// CHECK:       %[[T5:.*]] = unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T5:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T6:.*]] = llvm.insertvalue %[[T4]], %[[T5]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T7:.*]] = llvm.insertvalue %[[T4]], %[[T6]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T8:.*]] = llvm.insertvalue %[[T4]], %[[T7]][2] : !llvm.array<3 x vector<2xf32>>
-// CHECK:       %[[T9:.*]] = unrealized_conversion_cast %[[T1]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
+// CHECK:       %[[T9:.*]] = builtin.unrealized_conversion_cast %[[T1]] : vector<4x3x2xf32> to !llvm.array<4 x array<3 x vector<2xf32>>>
 // CHECK:       %[[T10:.*]] = llvm.insertvalue %[[T8]], %[[T9]][0] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T11:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
+// CHECK:       %[[T11:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
 // CHECK:       %[[T12:.*]] = llvm.extractvalue %[[T11]][1, 0] : !llvm.array<4 x array<1 x vector<2xf32>>>
-// CHECK:       %[[T13:.*]] = unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T13:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T14:.*]] = llvm.insertvalue %[[T12]], %[[T13]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T15:.*]] = llvm.insertvalue %[[T12]], %[[T14]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T16:.*]] = llvm.insertvalue %[[T12]], %[[T15]][2] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T17:.*]] = llvm.insertvalue %[[T16]], %[[T10]][1] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T18:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
+// CHECK:       %[[T18:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
 // CHECK:       %[[T19:.*]] = llvm.extractvalue %[[T18]][2, 0] : !llvm.array<4 x array<1 x vector<2xf32>>>
-// CHECK:       %[[T20:.*]] = unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T20:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T21:.*]] = llvm.insertvalue %[[T19]], %[[T20]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T22:.*]] = llvm.insertvalue %[[T19]], %[[T21]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T23:.*]] = llvm.insertvalue %[[T19]], %[[T22]][2] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T24:.*]] = llvm.insertvalue %[[T23]], %[[T17]][2] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T25:.*]] = unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
+// CHECK:       %[[T25:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4x1x2xf32> to !llvm.array<4 x array<1 x vector<2xf32>>>
 // CHECK:       %[[T26:.*]] = llvm.extractvalue %[[T25]][3, 0] : !llvm.array<4 x array<1 x vector<2xf32>>>
-// CHECK:       %[[T27:.*]] = unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
+// CHECK:       %[[T27:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<3x2xf32> to !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T28:.*]] = llvm.insertvalue %[[T26]], %[[T27]][0] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T29:.*]] = llvm.insertvalue %[[T26]], %[[T28]][1] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T30:.*]] = llvm.insertvalue %[[T26]], %[[T29]][2] : !llvm.array<3 x vector<2xf32>>
 // CHECK:       %[[T31:.*]] = llvm.insertvalue %[[T30]], %[[T24]][3] : !llvm.array<4 x array<3 x vector<2xf32>>>
-// CHECK:       %[[T32:.*]] = unrealized_conversion_cast %[[T31]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
+// CHECK:       %[[T32:.*]] = builtin.unrealized_conversion_cast %[[T31]] : !llvm.array<4 x array<3 x vector<2xf32>>> to vector<4x3x2xf32>
 // CHECK:       return %[[T32]] : vector<4x3x2xf32>
 
 // -----
 
 func @outerproduct(%arg0: vector<2xf32>, %arg1: vector<3xf32>) -> vector<2x3xf32> {
   %2 = vector.outerproduct %arg0, %arg1 : vector<2xf32>, vector<3xf32>
   return %2 : vector<2x3xf32>
 }
 // CHECK-LABEL: @outerproduct(
 // CHECK-SAME:  %[[A:.*]]: vector<2xf32>,
 // CHECK-SAME:  %[[B:.*]]: vector<3xf32>)
 // CHECK:       %[[T2:.*]] = constant dense<0.000000e+00> : vector<2x3xf32>
 // CHECK:       %[[T3:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T4:.*]] = llvm.extractelement %[[A]]{{\[}}%[[T3]] : i64] : vector<2xf32>
 // CHECK:       %[[T5:.*]] = splat %[[T4]] : vector<3xf32>
 // CHECK:       %[[T6:.*]] = mulf %[[T5]], %[[B]] : vector<3xf32>
-// CHECK:       %[[T7:.*]] = unrealized_conversion_cast %[[T2]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
+// CHECK:       %[[T7:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T8:.*]] = llvm.insertvalue %[[T6]], %[[T7]][0] : !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T9:.*]] = llvm.mlir.constant(1 : i64) : i64
 // CHECK:       %[[T10:.*]] = llvm.extractelement %[[A]]{{\[}}%[[T9]] : i64] : vector<2xf32>
 // CHECK:       %[[T11:.*]] = splat %[[T10]] : vector<3xf32>
 // CHECK:       %[[T12:.*]] = mulf %[[T11]], %[[B]] : vector<3xf32>
 // CHECK:       %[[T13:.*]] = llvm.insertvalue %[[T12]], %[[T8]][1] : !llvm.array<2 x vector<3xf32>>
-// CHECK:       %[[T14:.*]] = unrealized_conversion_cast %[[T13]] : !llvm.array<2 x vector<3xf32>> to vector<2x3xf32>
+// CHECK:       %[[T14:.*]] = builtin.unrealized_conversion_cast %[[T13]] : !llvm.array<2 x vector<3xf32>> to vector<2x3xf32>
 // CHECK:       return %[[T14]] : vector<2x3xf32>
 
 // -----
 
 func @outerproduct_index(%arg0: vector<2xindex>, %arg1: vector<3xindex>) -> vector<2x3xindex> {
   %2 = vector.outerproduct %arg0, %arg1 : vector<2xindex>, vector<3xindex>
   return %2 : vector<2x3xindex>
 }
 // CHECK-LABEL: @outerproduct_index(
 // CHECK-SAME:  %[[A:.*]]: vector<2xindex>,
 // CHECK-SAME:  %[[B:.*]]: vector<3xindex>)
 // CHECK:       %[[T0:.*]] = constant dense<0> : vector<2x3xindex>
-// CHECK:       %[[T1:.*]] = unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
+// CHECK:       %[[T1:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
 // CHECK:       %[[T2:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T3:.*]] = llvm.extractelement %[[T1]]{{\[}}%[[T2]] : i64] : vector<2xi64>
-// CHECK:       %[[T4:.*]] = unrealized_conversion_cast %[[T3]] : i64 to index
+// CHECK:       %[[T4:.*]] = builtin.unrealized_conversion_cast %[[T3]] : i64 to index
 // CHECK:       %[[T5:.*]] = splat %[[T4]] : vector<3xindex>
 // CHECK:       %[[T6:.*]] = muli %[[T5]], %[[B]] : vector<3xindex>
-// CHECK:       %[[T7:.*]] = unrealized_conversion_cast %[[T6]] : vector<3xindex> to vector<3xi64>
-// CHECK:       %[[T8:.*]] = unrealized_conversion_cast %[[T0]] : vector<2x3xindex> to !llvm.array<2 x vector<3xi64>>
+// CHECK:       %[[T7:.*]] = builtin.unrealized_conversion_cast %[[T6]] : vector<3xindex> to vector<3xi64>
+// CHECK:       %[[T8:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<2x3xindex> to !llvm.array<2 x vector<3xi64>>
 // CHECK:       %{{.*}} = llvm.insertvalue %[[T7]], %[[T8]][0] : !llvm.array<2 x vector<3xi64>>
 
 // -----
 
 func @outerproduct_add(%arg0: vector<2xf32>, %arg1: vector<3xf32>, %arg2: vector<2x3xf32>) -> vector<2x3xf32> {
   %2 = vector.outerproduct %arg0, %arg1, %arg2 : vector<2xf32>, vector<3xf32>
   return %2 : vector<2x3xf32>
 }
 // CHECK-LABEL: @outerproduct_add(
 // CHECK-SAME:  %[[A:.*]]: vector<2xf32>,
 // CHECK-SAME:  %[[B:.*]]: vector<3xf32>,
 // CHECK-SAME:  %[[C:.*]]: vector<2x3xf32>) -> vector<2x3xf32>
 // CHECK:       %[[T3:.*]] = constant dense<0.000000e+00> : vector<2x3xf32>
 // CHECK:       %[[T4:.*]] = llvm.mlir.constant(0 : i64) : i64
 // CHECK:       %[[T5:.*]] = llvm.extractelement %[[A]]{{\[}}%[[T4]] : i64] : vector<2xf32>
 // CHECK:       %[[T6:.*]] = splat %[[T5]] : vector<3xf32>
-// CHECK:       %[[T7:.*]] = unrealized_conversion_cast %[[C]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
+// CHECK:       %[[T7:.*]] = builtin.unrealized_conversion_cast %[[C]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T8:.*]] = llvm.extractvalue %[[T7]][0] : !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T9:.*]] = "llvm.intr.fmuladd"(%[[T6]], %[[B]], %[[T8]]) : (vector<3xf32>, vector<3xf32>, vector<3xf32>) -> vector<3xf32>
-// CHECK:       %[[T10:.*]] = unrealized_conversion_cast %[[T3]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
+// CHECK:       %[[T10:.*]] = builtin.unrealized_conversion_cast %[[T3]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T11:.*]] = llvm.insertvalue %[[T9]], %[[T10]][0] : !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T12:.*]] = llvm.mlir.constant(1 : i64) : i64
 // CHECK:       %[[T13:.*]] = llvm.extractelement %[[A]]{{\[}}%[[T12]] : i64] : vector<2xf32>
 // CHECK:       %[[T14:.*]] = splat %[[T13]] : vector<3xf32>
-// CHECK:       %[[T15:.*]] = unrealized_conversion_cast %[[C]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
+// CHECK:       %[[T15:.*]] = builtin.unrealized_conversion_cast %[[C]] : vector<2x3xf32> to !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T16:.*]] = llvm.extractvalue %[[T15]][1] : !llvm.array<2 x vector<3xf32>>
 // CHECK:       %[[T17:.*]] = "llvm.intr.fmuladd"(%[[T14]], %[[B]], %[[T16]]) : (vector<3xf32>, vector<3xf32>, vector<3xf32>) -> vector<3xf32>
 // CHECK:       %[[T18:.*]] = llvm.insertvalue %[[T17]], %[[T11]][1] : !llvm.array<2 x vector<3xf32>>
-// CHECK:       %[[T19:.*]] = unrealized_conversion_cast %[[T18]] : !llvm.array<2 x vector<3xf32>> to vector<2x3xf32>
+// CHECK:       %[[T19:.*]] = builtin.unrealized_conversion_cast %[[T18]] : !llvm.array<2 x vector<3xf32>> to vector<2x3xf32>
 // CHECK:       return %[[T19]] : vector<2x3xf32>
 
 // -----
 
 func @shuffle_1D_direct(%arg0: vector<2xf32>, %arg1: vector<2xf32>) -> vector<2xf32> {
   %1 = vector.shuffle %arg0, %arg1 [0, 1] : vector<2xf32>, vector<2xf32>
   return %1 : vector<2xf32>
 }
 // CHECK-LABEL: @shuffle_1D_direct(
 // CHECK-SAME: %[[A:.*]]: vector<2xf32>,
 // CHECK-SAME: %[[B:.*]]: vector<2xf32>)
 //       CHECK:   %[[s:.*]] = llvm.shufflevector %[[A]], %[[B]] [0, 1] : vector<2xf32>, vector<2xf32>
 //       CHECK:   return %[[s]] : vector<2xf32>
 
 // -----
 
 func @shuffle_1D_index_direct(%arg0: vector<2xindex>, %arg1: vector<2xindex>) -> vector<2xindex> {
   %1 = vector.shuffle %arg0, %arg1 [0, 1] : vector<2xindex>, vector<2xindex>
   return %1 : vector<2xindex>
 }
 // CHECK-LABEL: @shuffle_1D_index_direct(
 // CHECK-SAME: %[[A:.*]]: vector<2xindex>,
 // CHECK-SAME: %[[B:.*]]: vector<2xindex>)
-//       CHECK:   %[[T0:.*]] = unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
-//       CHECK:   %[[T1:.*]] = unrealized_conversion_cast %[[B]] : vector<2xindex> to vector<2xi64>
+//       CHECK:   %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2xindex> to vector<2xi64>
+//       CHECK:   %[[T1:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2xindex> to vector<2xi64>
 //       CHECK:   %[[T2:.*]] = llvm.shufflevector %[[T0]], %[[T1]] [0, 1] : vector<2xi64>, vector<2xi64>
-//       CHECK:   %[[T3:.*]] = unrealized_conversion_cast %[[T2]] : vector<2xi64> to vector<2xindex>
+//       CHECK:   %[[T3:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<2xi64> to vector<2xindex>
 //       CHECK:   return %[[T3]] : vector<2xindex>
 
 // -----
 
 func @shuffle_1D(%arg0: vector<2xf32>, %arg1: vector<3xf32>) -> vector<5xf32> {
   %1 = vector.shuffle %arg0, %arg1 [4, 3, 2, 1, 0] : vector<2xf32>, vector<3xf32>
   return %1 : vector<5xf32>
 }

 
 func @shuffle_2D(%a: vector<1x4xf32>, %b: vector<2x4xf32>) -> vector<3x4xf32> {
   %1 = vector.shuffle %a, %b[1, 0, 2] : vector<1x4xf32>, vector<2x4xf32>
   return %1 : vector<3x4xf32>
 }
 // CHECK-LABEL: @shuffle_2D(
 // CHECK-SAME: %[[A:.*]]: vector<1x4xf32>,
 // CHECK-SAME: %[[B:.*]]: vector<2x4xf32>)
-//       CHECK:   %[[VAL_0:.*]] = unrealized_conversion_cast %[[A]] : vector<1x4xf32> to !llvm.array<1 x vector<4xf32>>
-//       CHECK:   %[[VAL_1:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+//       CHECK:   %[[VAL_0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<1x4xf32> to !llvm.array<1 x vector<4xf32>>
+//       CHECK:   %[[VAL_1:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
 //       CHECK:   %[[u0:.*]] = llvm.mlir.undef : !llvm.array<3 x vector<4xf32>>
 //       CHECK:   %[[e1:.*]] = llvm.extractvalue %[[VAL_1]][0] : !llvm.array<2 x vector<4xf32>>
 //       CHECK:   %[[i1:.*]] = llvm.insertvalue %[[e1]], %[[u0]][0] : !llvm.array<3 x vector<4xf32>>
 //       CHECK:   %[[e2:.*]] = llvm.extractvalue %[[VAL_0]][0] : !llvm.array<1 x vector<4xf32>>
 //       CHECK:   %[[i2:.*]] = llvm.insertvalue %[[e2]], %[[i1]][1] : !llvm.array<3 x vector<4xf32>>
 //       CHECK:   %[[e3:.*]] = llvm.extractvalue %[[VAL_1]][1] : !llvm.array<2 x vector<4xf32>>
 //       CHECK:   %[[i3:.*]] = llvm.insertvalue %[[e3]], %[[i2]][2] : !llvm.array<3 x vector<4xf32>>
-//       CHECK:   %[[VAL_3:.*]] = unrealized_conversion_cast %[[i3]] : !llvm.array<3 x vector<4xf32>> to vector<3x4xf32>
+//       CHECK:   %[[VAL_3:.*]] = builtin.unrealized_conversion_cast %[[i3]] : !llvm.array<3 x vector<4xf32>> to vector<3x4xf32>
 //       CHECK:   return %[[VAL_3]] : vector<3x4xf32>
 
 // -----
 
 func @extract_element(%arg0: vector<16xf32>) -> f32 {
   %0 = constant 15 : i32
   %1 = vector.extractelement %arg0[%0 : i32]: vector<16xf32>
   return %1 : f32

 // -----
 
 func @extract_index_element_from_vec_1d(%arg0: vector<16xindex>) -> index {
   %0 = vector.extract %arg0[15]: vector<16xindex>
   return %0 : index
 }
 // CHECK-LABEL: @extract_index_element_from_vec_1d(
 // CHECK-SAME: %[[A:.*]]: vector<16xindex>)
-//       CHECK:   %[[T0:.*]] = unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
+//       CHECK:   %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
 //       CHECK:   %[[T1:.*]] = llvm.mlir.constant(15 : i64) : i64
 //       CHECK:   %[[T2:.*]] = llvm.extractelement %[[T0]][%[[T1]] : i64] : vector<16xi64>
-//       CHECK:   %[[T3:.*]] = unrealized_conversion_cast %[[T2]] : i64 to index
+//       CHECK:   %[[T3:.*]] = builtin.unrealized_conversion_cast %[[T2]] : i64 to index
 //       CHECK:   return %[[T3]] : index
 
 // -----
 
 func @extract_vec_2d_from_vec_3d(%arg0: vector<4x3x16xf32>) -> vector<3x16xf32> {
   %0 = vector.extract %arg0[0]: vector<4x3x16xf32>
   return %0 : vector<3x16xf32>
 }

 
 func @insert_index_element_into_vec_1d(%arg0: index, %arg1: vector<4xindex>) -> vector<4xindex> {
   %0 = vector.insert %arg0, %arg1[3] : index into vector<4xindex>
   return %0 : vector<4xindex>
 }
 // CHECK-LABEL: @insert_index_element_into_vec_1d(
 // CHECK-SAME: %[[A:.*]]: index,
 // CHECK-SAME: %[[B:.*]]: vector<4xindex>)
-//       CHECK:   %[[T0:.*]] = unrealized_conversion_cast %[[A]] : index to i64
-//       CHECK:   %[[T1:.*]] = unrealized_conversion_cast %[[B]] : vector<4xindex> to vector<4xi64>
+//       CHECK:   %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : index to i64
+//       CHECK:   %[[T1:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<4xindex> to vector<4xi64>
 //       CHECK:   %[[T3:.*]] = llvm.mlir.constant(3 : i64) : i64
 //       CHECK:   %[[T4:.*]] = llvm.insertelement %[[T0]], %[[T1]][%[[T3]] : i64] : vector<4xi64>
-//       CHECK:   %[[T5:.*]] = unrealized_conversion_cast %[[T4]] : vector<4xi64> to vector<4xindex>
+//       CHECK:   %[[T5:.*]] = builtin.unrealized_conversion_cast %[[T4]] : vector<4xi64> to vector<4xindex>
 //       CHECK:   return %[[T5]] : vector<4xindex>
 
 // -----
 
 func @insert_vec_2d_into_vec_3d(%arg0: vector<8x16xf32>, %arg1: vector<4x8x16xf32>) -> vector<4x8x16xf32> {
   %0 = vector.insert %arg0, %arg1[3] : vector<8x16xf32> into vector<4x8x16xf32>
   return %0 : vector<4x8x16xf32>
 }

 // -----
 
 func @vector_index_type_cast(%arg0: memref<8x8x8xindex>) -> memref<vector<8x8x8xindex>> {
   %0 = vector.type_cast %arg0: memref<8x8x8xindex> to memref<vector<8x8x8xindex>>
   return %0 : memref<vector<8x8x8xindex>>
 }
 // CHECK-LABEL: @vector_index_type_cast(
 // CHECK-SAME: %[[A:.*]]: memref<8x8x8xindex>)
-//       CHECK:   %{{.*}} = unrealized_conversion_cast %[[A]] : memref<8x8x8xindex> to !llvm.struct<(ptr<i64>, ptr<i64>, i64, array<3 x i64>, array<3 x i64>)>
+//       CHECK:   %{{.*}} = builtin.unrealized_conversion_cast %[[A]] : memref<8x8x8xindex> to !llvm.struct<(ptr<i64>, ptr<i64>, i64, array<3 x i64>, array<3 x i64>)>
 
-//       CHECK:   %{{.*}} = unrealized_conversion_cast %{{.*}} : !llvm.struct<(ptr<array<8 x array<8 x vector<8xi64>>>>, ptr<array<8 x array<8 x vector<8xi64>>>>, i64)> to memref<vector<8x8x8xindex>>
+//       CHECK:   %{{.*}} = builtin.unrealized_conversion_cast %{{.*}} : !llvm.struct<(ptr<array<8 x array<8 x vector<8xi64>>>>, ptr<array<8 x array<8 x vector<8xi64>>>>, i64)> to memref<vector<8x8x8xindex>>
 
 // -----
 
 func @vector_type_cast_non_zero_addrspace(%arg0: memref<8x8x8xf32, 3>) -> memref<vector<8x8x8xf32>, 3> {
   %0 = vector.type_cast %arg0: memref<8x8x8xf32, 3> to memref<vector<8x8x8xf32>, 3>
   return %0 : memref<vector<8x8x8xf32>, 3>
 }
 // CHECK-LABEL: @vector_type_cast_non_zero_addrspace

 // -----
 
 func @vector_print_scalar_si4(%arg0: si4) {
   vector.print %arg0 : si4
   return
 }
 // CHECK-LABEL: @vector_print_scalar_si4(
 // CHECK-SAME: %[[A:.*]]: si4)
-//       CHECK: %[[C:.*]] = unrealized_conversion_cast %[[A]] : si4 to i4
+//       CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : si4 to i4
 //       CHECK: %[[S:.*]] = sexti %[[C]] : i4 to i64
 //       CHECK: llvm.call @printI64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_ui4(%arg0: ui4) {
   vector.print %arg0 : ui4
   return
 }
 // CHECK-LABEL: @vector_print_scalar_ui4(
 // CHECK-SAME: %[[A:.*]]: ui4)
-//       CHECK: %[[C:.*]] = unrealized_conversion_cast %[[A]] : ui4 to i4
+//       CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : ui4 to i4
 //       CHECK: %[[S:.*]] = zexti %[[C]] : i4 to i64
 //       CHECK: llvm.call @printU64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_i32(%arg0: i32) {
   vector.print %arg0 : i32
   return
 }
 // CHECK-LABEL: @vector_print_scalar_i32(
 // CHECK-SAME: %[[A:.*]]: i32)
 //       CHECK: %[[S:.*]] = sexti %[[A]] : i32 to i64
 //       CHECK: llvm.call @printI64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_ui32(%arg0: ui32) {
   vector.print %arg0 : ui32
   return
 }
 // CHECK-LABEL: @vector_print_scalar_ui32(
 // CHECK-SAME: %[[A:.*]]: ui32)
-//       CHECK: %[[C:.*]] = unrealized_conversion_cast %[[A]] : ui32 to i32
+//       CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : ui32 to i32
 //       CHECK: %[[S:.*]] = zexti %[[C]] : i32 to i64
 //       CHECK: llvm.call @printU64(%[[S]]) : (i64) -> ()
 
 // -----
 
 func @vector_print_scalar_i40(%arg0: i40) {
   vector.print %arg0 : i40
   return
 }
 // CHECK-LABEL: @vector_print_scalar_i40(
 // CHECK-SAME: %[[A:.*]]: i40)
 //       CHECK: %[[S:.*]] = sexti %[[A]] : i40 to i64
 //       CHECK: llvm.call @printI64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_si40(%arg0: si40) {
   vector.print %arg0 : si40
   return
 }
 // CHECK-LABEL: @vector_print_scalar_si40(
 // CHECK-SAME: %[[A:.*]]: si40)
-//       CHECK: %[[C:.*]] = unrealized_conversion_cast %[[A]] : si40 to i40
+//       CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : si40 to i40
 //       CHECK: %[[S:.*]] = sexti %[[C]] : i40 to i64
 //       CHECK: llvm.call @printI64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_ui40(%arg0: ui40) {
   vector.print %arg0 : ui40
   return
 }
 // CHECK-LABEL: @vector_print_scalar_ui40(
 // CHECK-SAME: %[[A:.*]]: ui40)
-//       CHECK: %[[C:.*]] = unrealized_conversion_cast %[[A]] : ui40 to i40
+//       CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : ui40 to i40
 //       CHECK: %[[S:.*]] = zexti %[[C]] : i40 to i64
 //       CHECK: llvm.call @printU64(%[[S]]) : (i64) -> ()
 //       CHECK: llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_i64(%arg0: i64) {
   vector.print %arg0 : i64
   return
 }
 // CHECK-LABEL: @vector_print_scalar_i64(
 // CHECK-SAME: %[[A:.*]]: i64)
 //       CHECK:    llvm.call @printI64(%[[A]]) : (i64) -> ()
 //       CHECK:    llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_ui64(%arg0: ui64) {
   vector.print %arg0 : ui64
   return
 }
 // CHECK-LABEL: @vector_print_scalar_ui64(
 // CHECK-SAME: %[[A:.*]]: ui64)
-//       CHECK:    %[[C:.*]] = unrealized_conversion_cast %[[A]] : ui64 to i64
+//       CHECK:    %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : ui64 to i64
 //       CHECK:    llvm.call @printU64(%[[C]]) : (i64) -> ()
 //       CHECK:    llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_index(%arg0: index) {
   vector.print %arg0 : index
   return
 }
 // CHECK-LABEL: @vector_print_scalar_index(
 // CHECK-SAME: %[[A:.*]]: index)
-//       CHECK:    %[[C:.*]] = unrealized_conversion_cast %[[A]] : index to i64
+//       CHECK:    %[[C:.*]] = builtin.unrealized_conversion_cast %[[A]] : index to i64
 //       CHECK:    llvm.call @printU64(%[[C]]) : (i64) -> ()
 //       CHECK:    llvm.call @printNewline() : () -> ()
 
 // -----
 
 func @vector_print_scalar_f32(%arg0: f32) {
   vector.print %arg0 : f32
   return

 // -----
 
 func @vector_print_vector(%arg0: vector<2x2xf32>) {
   vector.print %arg0 : vector<2x2xf32>
   return
 }
 // CHECK-LABEL: @vector_print_vector(
 // CHECK-SAME: %[[A:.*]]: vector<2x2xf32>)
-//       CHECK:    %[[VAL_1:.*]] = unrealized_conversion_cast %[[A]] : vector<2x2xf32> to !llvm.array<2 x vector<2xf32>>
+//       CHECK:    %[[VAL_1:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2x2xf32> to !llvm.array<2 x vector<2xf32>>
 //       CHECK:    llvm.call @printOpen() : () -> ()
 //       CHECK:    %[[x0:.*]] = llvm.extractvalue %[[VAL_1]][0] : !llvm.array<2 x vector<2xf32>>
 //       CHECK:    llvm.call @printOpen() : () -> ()
 //       CHECK:    %[[x1:.*]] = llvm.mlir.constant(0 : index) : i64
 //       CHECK:    %[[x2:.*]] = llvm.extractelement %[[x0]][%[[x1]] : i64] : vector<2xf32>
 //       CHECK:    llvm.call @printF32(%[[x2]]) : (f32) -> ()
 //       CHECK:    llvm.call @printComma() : () -> ()
 //       CHECK:    %[[x3:.*]] = llvm.mlir.constant(1 : index) : i64

 // -----
 
 func @extract_strided_index_slice1(%arg0: vector<4xindex>) -> vector<2xindex> {
   %0 = vector.extract_strided_slice %arg0 {offsets = [2], sizes = [2], strides = [1]} : vector<4xindex> to vector<2xindex>
   return %0 : vector<2xindex>
 }
 // CHECK-LABEL: @extract_strided_index_slice1(
 //  CHECK-SAME:    %[[A:.*]]: vector<4xindex>)
-//       CHECK:    %[[T0:.*]] = unrealized_conversion_cast %[[A]] : vector<4xindex> to vector<4xi64>
-//       CHECK:    %[[T1:.*]] = unrealized_conversion_cast %[[A]] : vector<4xindex> to vector<4xi64>
+//       CHECK:    %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4xindex> to vector<4xi64>
+//       CHECK:    %[[T1:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<4xindex> to vector<4xi64>
 //       CHECK:    %[[T2:.*]] = llvm.shufflevector %[[T0]], %[[T1]] [2, 3] : vector<4xi64>, vector<4xi64>
-//       CHECK:    %[[T3:.*]] = unrealized_conversion_cast %[[T2]] : vector<2xi64> to vector<2xindex>
+//       CHECK:    %[[T3:.*]] = builtin.unrealized_conversion_cast %[[T2]] : vector<2xi64> to vector<2xindex>
 //       CHECK:    return %[[T3]] : vector<2xindex>
 
 // -----
 
 func @extract_strided_slice2(%arg0: vector<4x8xf32>) -> vector<2x8xf32> {
   %0 = vector.extract_strided_slice %arg0 {offsets = [2], sizes = [2], strides = [1]} : vector<4x8xf32> to vector<2x8xf32>
   return %0 : vector<2x8xf32>
 }
 // CHECK-LABEL: @extract_strided_slice2(
 //  CHECK-SAME:    %[[ARG:.*]]: vector<4x8xf32>)
-//       CHECK:    %[[A:.*]] = unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
+//       CHECK:    %[[A:.*]] = builtin.unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T0:.*]] = llvm.mlir.undef : !llvm.array<2 x vector<8xf32>>
 //       CHECK:    %[[T1:.*]] = llvm.extractvalue %[[A]][2] : !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T2:.*]] = llvm.insertvalue %[[T1]], %[[T0]][0] : !llvm.array<2 x vector<8xf32>>
 //       CHECK:    %[[T3:.*]] = llvm.extractvalue %[[A]][3] : !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T4:.*]] = llvm.insertvalue %[[T3]], %[[T2]][1] : !llvm.array<2 x vector<8xf32>>
-//       CHECK:    %[[T5:.*]] = unrealized_conversion_cast %[[T4]] : !llvm.array<2 x vector<8xf32>> to vector<2x8xf32>
+//       CHECK:    %[[T5:.*]] = builtin.unrealized_conversion_cast %[[T4]] : !llvm.array<2 x vector<8xf32>> to vector<2x8xf32>
 //       CHECK:    return %[[T5]]
 
 // -----
 
 func @extract_strided_slice3(%arg0: vector<4x8xf32>) -> vector<2x2xf32> {
   %0 = vector.extract_strided_slice %arg0 {offsets = [2, 2], sizes = [2, 2], strides = [1, 1]} : vector<4x8xf32> to vector<2x2xf32>
   return %0 : vector<2x2xf32>
 }
 // CHECK-LABEL: @extract_strided_slice3(
 //  CHECK-SAME:    %[[ARG:.*]]: vector<4x8xf32>)
 //       CHECK:    %[[VAL_1:.*]] = constant 0.000000e+00 : f32
 //       CHECK:    %[[VAL_2:.*]] = splat %[[VAL_1]] : vector<2x2xf32>
-//       CHECK:    %[[A:.*]] = unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
+//       CHECK:    %[[A:.*]] = builtin.unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T2:.*]] = llvm.extractvalue %[[A]][2] : !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T3:.*]] = llvm.shufflevector %[[T2]], %[[T2]] [2, 3] : vector<8xf32>, vector<8xf32>
-//       CHECK:    %[[VAL_6:.*]] = unrealized_conversion_cast %[[VAL_2]] : vector<2x2xf32> to !llvm.array<2 x vector<2xf32>>
+//       CHECK:    %[[VAL_6:.*]] = builtin.unrealized_conversion_cast %[[VAL_2]] : vector<2x2xf32> to !llvm.array<2 x vector<2xf32>>
 //       CHECK:    %[[T4:.*]] = llvm.insertvalue %[[T3]], %[[VAL_6]][0] : !llvm.array<2 x vector<2xf32>>
-//       CHECK:    %[[A:.*]] = unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
+//       CHECK:    %[[A:.*]] = builtin.unrealized_conversion_cast %[[ARG]] : vector<4x8xf32> to !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T5:.*]] = llvm.extractvalue %[[A]][3] : !llvm.array<4 x vector<8xf32>>
 //       CHECK:    %[[T6:.*]] = llvm.shufflevector %[[T5]], %[[T5]] [2, 3] : vector<8xf32>, vector<8xf32>
 //       CHECK:    %[[T7:.*]] = llvm.insertvalue %[[T6]], %[[T4]][1] : !llvm.array<2 x vector<2xf32>>
-//       CHECK:    %[[VAL_12:.*]] = unrealized_conversion_cast %[[T7]] : !llvm.array<2 x vector<2xf32>> to vector<2x2xf32>
+//       CHECK:    %[[VAL_12:.*]] = builtin.unrealized_conversion_cast %[[T7]] : !llvm.array<2 x vector<2xf32>> to vector<2x2xf32>
 //       CHECK:    return %[[VAL_12]] : vector<2x2xf32>
 
 // -----
 
 func @insert_strided_slice1(%b: vector<4x4xf32>, %c: vector<4x4x4xf32>) -> vector<4x4x4xf32> {
   %0 = vector.insert_strided_slice %b, %c {offsets = [2, 0, 0], strides = [1, 1]} : vector<4x4xf32> into vector<4x4x4xf32>
   return %0 : vector<4x4x4xf32>
 }

 func @insert_strided_slice3(%arg0: vector<2x4xf32>, %arg1: vector<16x4x8xf32>) -> vector<16x4x8xf32> {
   %0 = vector.insert_strided_slice %arg0, %arg1 {offsets = [0, 0, 2], strides = [1, 1]}:
         vector<2x4xf32> into vector<16x4x8xf32>
   return %0 : vector<16x4x8xf32>
 }
 // CHECK-LABEL: @insert_strided_slice3(
 // CHECK-SAME: %[[A:.*]]: vector<2x4xf32>,
 // CHECK-SAME: %[[B:.*]]: vector<16x4x8xf32>)
-//      CHECK: %[[s2:.*]] = unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
+//      CHECK: %[[s2:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s3:.*]] = llvm.extractvalue %[[s2]][0] : !llvm.array<16 x array<4 x vector<8xf32>>>
-//      CHECK: %[[s4:.*]] = unrealized_conversion_cast %[[A]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+//      CHECK: %[[s4:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
 //      CHECK: %[[s5:.*]] = llvm.extractvalue %[[s4]][0] : !llvm.array<2 x vector<4xf32>>
-//      CHECK: %[[s6:.*]] = unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
+//      CHECK: %[[s6:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s7:.*]] = llvm.extractvalue %[[s6]][0, 0] : !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s8:.*]] = constant 0 : index
-//      CHECK: %[[s9:.*]] = unrealized_conversion_cast %[[s8]] : index to i64
+//      CHECK: %[[s9:.*]] = builtin.unrealized_conversion_cast %[[s8]] : index to i64
 //      CHECK: %[[s10:.*]] = llvm.extractelement %[[s5]]{{\[}}%[[s9]] : i64] : vector<4xf32>
 //      CHECK: %[[s11:.*]] = constant 2 : index
-//      CHECK: %[[s12:.*]] = unrealized_conversion_cast %[[s11]] : index to i64
+//      CHECK: %[[s12:.*]] = builtin.unrealized_conversion_cast %[[s11]] : index to i64
 //      CHECK: %[[s13:.*]] = llvm.insertelement %[[s10]], %[[s7]]{{\[}}%[[s12]] : i64] : vector<8xf32>
 //      CHECK: %[[s14:.*]] = constant 1 : index
-//      CHECK: %[[s15:.*]] = unrealized_conversion_cast %[[s14]] : index to i64
+//      CHECK: %[[s15:.*]] = builtin.unrealized_conversion_cast %[[s14]] : index to i64
 //      CHECK: %[[s16:.*]] = llvm.extractelement %[[s5]]{{\[}}%[[s15]] : i64] : vector<4xf32>
 //      CHECK: %[[s17:.*]] = constant 3 : index
-//      CHECK: %[[s18:.*]] = unrealized_conversion_cast %[[s17]] : index to i64
+//      CHECK: %[[s18:.*]] = builtin.unrealized_conversion_cast %[[s17]] : index to i64
 //      CHECK: %[[s19:.*]] = llvm.insertelement %[[s16]], %[[s13]]{{\[}}%[[s18]] : i64] : vector<8xf32>
 //      CHECK: %[[s20:.*]] = constant 2 : index
-//      CHECK: %[[s21:.*]] = unrealized_conversion_cast %[[s20]] : index to i64
+//      CHECK: %[[s21:.*]] = builtin.unrealized_conversion_cast %[[s20]] : index to i64
 //      CHECK: %[[s22:.*]] = llvm.extractelement %[[s5]]{{\[}}%[[s21]] : i64] : vector<4xf32>
 //      CHECK: %[[s23:.*]] = constant 4 : index
-//      CHECK: %[[s24:.*]] = unrealized_conversion_cast %[[s23]] : index to i64
+//      CHECK: %[[s24:.*]] = builtin.unrealized_conversion_cast %[[s23]] : index to i64
 //      CHECK: %[[s25:.*]] = llvm.insertelement %[[s22]], %[[s19]]{{\[}}%[[s24]] : i64] : vector<8xf32>
 //      CHECK: %[[s26:.*]] = constant 3 : index
-//      CHECK: %[[s27:.*]] = unrealized_conversion_cast %[[s26]] : index to i64
+//      CHECK: %[[s27:.*]] = builtin.unrealized_conversion_cast %[[s26]] : index to i64
 //      CHECK: %[[s28:.*]] = llvm.extractelement %[[s5]]{{\[}}%[[s27]] : i64] : vector<4xf32>
 //      CHECK: %[[s29:.*]] = constant 5 : index
-//      CHECK: %[[s30:.*]] = unrealized_conversion_cast %[[s29]] : index to i64
+//      CHECK: %[[s30:.*]] = builtin.unrealized_conversion_cast %[[s29]] : index to i64
 //      CHECK: %[[s31:.*]] = llvm.insertelement %[[s28]], %[[s25]]{{\[}}%[[s30]] : i64] : vector<8xf32>
 //      CHECK: %[[s32:.*]] = llvm.insertvalue %[[s31]], %[[s3]][0] : !llvm.array<4 x vector<8xf32>>
-//      CHECK: %[[s33:.*]] = unrealized_conversion_cast %[[A]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+//      CHECK: %[[s33:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
 //      CHECK: %[[s34:.*]] = llvm.extractvalue %[[s33]][1] : !llvm.array<2 x vector<4xf32>>
-//      CHECK: %[[s35:.*]] = unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
+//      CHECK: %[[s35:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s36:.*]] = llvm.extractvalue %[[s35]][0, 1] : !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s37:.*]] = constant 0 : index
-//      CHECK: %[[s38:.*]] = unrealized_conversion_cast %[[s37]] : index to i64
+//      CHECK: %[[s38:.*]] = builtin.unrealized_conversion_cast %[[s37]] : index to i64
 //      CHECK: %[[s39:.*]] = llvm.extractelement %[[s34]]{{\[}}%[[s38]] : i64] : vector<4xf32>
 //      CHECK: %[[s40:.*]] = constant 2 : index
-//      CHECK: %[[s41:.*]] = unrealized_conversion_cast %[[s40]] : index to i64
+//      CHECK: %[[s41:.*]] = builtin.unrealized_conversion_cast %[[s40]] : index to i64
 //      CHECK: %[[s42:.*]] = llvm.insertelement %[[s39]], %[[s36]]{{\[}}%[[s41]] : i64] : vector<8xf32>
 //      CHECK: %[[s43:.*]] = constant 1 : index
-//      CHECK: %[[s44:.*]] = unrealized_conversion_cast %[[s43]] : index to i64
+//      CHECK: %[[s44:.*]] = builtin.unrealized_conversion_cast %[[s43]] : index to i64
 //      CHECK: %[[s45:.*]] = llvm.extractelement %[[s34]]{{\[}}%[[s44]] : i64] : vector<4xf32>
 //      CHECK: %[[s46:.*]] = constant 3 : index
-//      CHECK: %[[s47:.*]] = unrealized_conversion_cast %[[s46]] : index to i64
+//      CHECK: %[[s47:.*]] = builtin.unrealized_conversion_cast %[[s46]] : index to i64
 //      CHECK: %[[s48:.*]] = llvm.insertelement %[[s45]], %[[s42]]{{\[}}%[[s47]] : i64] : vector<8xf32>
 //      CHECK: %[[s49:.*]] = constant 2 : index
-//      CHECK: %[[s50:.*]] = unrealized_conversion_cast %[[s49]] : index to i64
+//      CHECK: %[[s50:.*]] = builtin.unrealized_conversion_cast %[[s49]] : index to i64
 //      CHECK: %[[s51:.*]] = llvm.extractelement %[[s34]]{{\[}}%[[s50]] : i64] : vector<4xf32>
 //      CHECK: %[[s52:.*]] = constant 4 : index
-//      CHECK: %[[s53:.*]] = unrealized_conversion_cast %[[s52]] : index to i64
+//      CHECK: %[[s53:.*]] = builtin.unrealized_conversion_cast %[[s52]] : index to i64
 //      CHECK: %[[s54:.*]] = llvm.insertelement %[[s51]], %[[s48]]{{\[}}%[[s53]] : i64] : vector<8xf32>
 //      CHECK: %[[s55:.*]] = constant 3 : index
-//      CHECK: %[[s56:.*]] = unrealized_conversion_cast %[[s55]] : index to i64
+//      CHECK: %[[s56:.*]] = builtin.unrealized_conversion_cast %[[s55]] : index to i64
 //      CHECK: %[[s57:.*]] = llvm.extractelement %[[s34]]{{\[}}%[[s56]] : i64] : vector<4xf32>
 //      CHECK: %[[s58:.*]] = constant 5 : index
-//      CHECK: %[[s59:.*]] = unrealized_conversion_cast %[[s58]] : index to i64
+//      CHECK: %[[s59:.*]] = builtin.unrealized_conversion_cast %[[s58]] : index to i64
 //      CHECK: %[[s60:.*]] = llvm.insertelement %[[s57]], %[[s54]]{{\[}}%[[s59]] : i64] : vector<8xf32>
 //      CHECK: %[[s61:.*]] = llvm.insertvalue %[[s60]], %[[s32]][1] : !llvm.array<4 x vector<8xf32>>
-//      CHECK: %[[s62:.*]] = unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
+//      CHECK: %[[s62:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<16x4x8xf32> to !llvm.array<16 x array<4 x vector<8xf32>>>
 //      CHECK: %[[s63:.*]] = llvm.insertvalue %[[s61]], %[[s62]][0] : !llvm.array<16 x array<4 x vector<8xf32>>>
-//      CHECK: %[[s64:.*]] = unrealized_conversion_cast %[[s63]] : !llvm.array<16 x array<4 x vector<8xf32>>> to vector<16x4x8xf32>
+//      CHECK: %[[s64:.*]] = builtin.unrealized_conversion_cast %[[s63]] : !llvm.array<16 x array<4 x vector<8xf32>>> to vector<16x4x8xf32>
 //      CHECK: return %[[s64]] : vector<16x4x8xf32>
 
 // -----
 
 func @vector_fma(%a: vector<8xf32>, %b: vector<2x4xf32>) -> (vector<8xf32>, vector<2x4xf32>) {
   // CHECK-LABEL: @vector_fma
   //  CHECK-SAME: %[[A:.*]]: vector<8xf32>
   //  CHECK-SAME: %[[B:.*]]: vector<2x4xf32>
   //       CHECK: "llvm.intr.fmuladd"
   //  CHECK-SAME:   (vector<8xf32>, vector<8xf32>, vector<8xf32>) -> vector<8xf32>
   %0 = vector.fma %a, %a, %a : vector<8xf32>
 
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b00:.*]] = llvm.extractvalue %[[BL]][0] : !llvm.array<2 x vector<4xf32>>
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b01:.*]] = llvm.extractvalue %[[BL]][0] : !llvm.array<2 x vector<4xf32>>
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b02:.*]] = llvm.extractvalue %[[BL]][0] : !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[B0:.*]] = "llvm.intr.fmuladd"(%[[b00]], %[[b01]], %[[b02]]) :
   //  CHECK-SAME: (vector<4xf32>, vector<4xf32>, vector<4xf32>) -> vector<4xf32>
   //       CHECK: llvm.insertvalue %[[B0]], {{.*}}[0] : !llvm.array<2 x vector<4xf32>>
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b10:.*]] = llvm.extractvalue %[[BL]][1] : !llvm.array<2 x vector<4xf32>>
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b11:.*]] = llvm.extractvalue %[[BL]][1] : !llvm.array<2 x vector<4xf32>>
-  //       CHECK: %[[BL:.*]] = unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
+  //       CHECK: %[[BL:.*]] = builtin.unrealized_conversion_cast %[[B]] : vector<2x4xf32> to !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[b12:.*]] = llvm.extractvalue %[[BL]][1] : !llvm.array<2 x vector<4xf32>>
   //       CHECK: %[[B1:.*]] = "llvm.intr.fmuladd"(%[[b10]], %[[b11]], %[[b12]]) :
   //  CHECK-SAME: (vector<4xf32>, vector<4xf32>, vector<4xf32>) -> vector<4xf32>
   //       CHECK: llvm.insertvalue %[[B1]], {{.*}}[1] : !llvm.array<2 x vector<4xf32>>
   %1 = vector.fma %b, %b, %b : vector<2x4xf32>
 
   return %0, %1: vector<8xf32>, vector<2x4xf32>
 }

 // -----
 
 func @reduce_index(%arg0: vector<16xindex>) -> index {
   %0 = vector.reduction "add", %arg0 : vector<16xindex> into index
   return %0 : index
 }
 // CHECK-LABEL: @reduce_index(
 // CHECK-SAME: %[[A:.*]]: vector<16xindex>)
-//      CHECK: %[[T0:.*]] = unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
+//      CHECK: %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
 //      CHECK: %[[T1:.*]] = "llvm.intr.vector.reduce.add"(%[[T0]])
-//      CHECK: %[[T2:.*]] = unrealized_conversion_cast %[[T1]] : i64 to index
+//      CHECK: %[[T2:.*]] = builtin.unrealized_conversion_cast %[[T1]] : i64 to index
 //      CHECK: return %[[T2]] : index
 
 //                          4x16                16x3               4x3
 // -----
 
 func @matrix_ops(%A: vector<64xf64>, %B: vector<48xf64>) -> vector<12xf64> {
   %C = vector.matrix_multiply %A, %B
     { lhs_rows = 4: i32, lhs_columns = 16: i32 , rhs_columns = 3: i32 } :
vector.transfer_write %f, %A[%base]
       {permutation_map = affine_map<(d0) -> (d0)>} :
     vector<17xindex>, memref<?xindex>
   return %f: vector<17xindex>
 }
 // CHECK-LABEL: func @transfer_read_index_1d
 //  CHECK-SAME: %[[BASE:[a-zA-Z0-9]*]]: index) -> vector<17xindex>
 //       CHECK: %[[C7:.*]] = constant 7 : index
 //       CHECK: %[[SPLAT:.*]] = splat %[[C7]] : vector<17xindex>
-//       CHECK: %{{.*}} = unrealized_conversion_cast %[[SPLAT]] : vector<17xindex> to vector<17xi64>
+//       CHECK: %{{.*}} = builtin.unrealized_conversion_cast %[[SPLAT]] : vector<17xindex> to vector<17xi64>
 
 //       CHECK: %[[loaded:.*]] = llvm.intr.masked.load %{{.*}}, %{{.*}}, %{{.*}} {alignment = 8 : i32} :
 //  CHECK-SAME: (!llvm.ptr<vector<17xi64>>, vector<17xi1>, vector<17xi64>) -> vector<17xi64>
 
 //       CHECK: llvm.intr.masked.store %[[loaded]], %{{.*}}, %{{.*}} {alignment = 8 : i32} :
 //  CHECK-SAME: vector<17xi64>, vector<17xi1> into !llvm.ptr<vector<17xi64>>
 
 // -----

 func @genbool_2d() -> vector<4x4xi1> {
   %v = vector.constant_mask [2, 2] : vector<4x4xi1>
   return %v: vector<4x4xi1>
 }
 
 // CHECK-LABEL: func @genbool_2d
 // CHECK: %[[VAL_0:.*]] = constant dense<[true, true, false, false]> : vector<4xi1>
 // CHECK: %[[VAL_1:.*]] = constant dense<false> : vector<4x4xi1>
-// CHECK: %[[VAL_2:.*]] = unrealized_conversion_cast %[[VAL_1]] : vector<4x4xi1> to !llvm.array<4 x vector<4xi1>>
+// CHECK: %[[VAL_2:.*]] = builtin.unrealized_conversion_cast %[[VAL_1]] : vector<4x4xi1> to !llvm.array<4 x vector<4xi1>>
 // CHECK: %[[VAL_3:.*]] = llvm.insertvalue %[[VAL_0]], %[[VAL_2]][0] : !llvm.array<4 x vector<4xi1>>
 // CHECK: %[[VAL_4:.*]] = llvm.insertvalue %[[VAL_0]], %[[VAL_3]][1] : !llvm.array<4 x vector<4xi1>>
-// CHECK: %[[VAL_5:.*]] = unrealized_conversion_cast %[[VAL_4]] : !llvm.array<4 x vector<4xi1>> to vector<4x4xi1>
+// CHECK: %[[VAL_5:.*]] = builtin.unrealized_conversion_cast %[[VAL_4]] : !llvm.array<4 x vector<4xi1>> to vector<4x4xi1>
 // CHECK: return %[[VAL_5]] : vector<4x4xi1>
 
 // -----
 
 func @flat_transpose(%arg0: vector<16xf32>) -> vector<16xf32> {
   %0 = vector.flat_transpose %arg0 { rows = 4: i32, columns = 4: i32 }
      : vector<16xf32> -> vector<16xf32>
   return %0 : vector<16xf32>

 
 func @flat_transpose_index(%arg0: vector<16xindex>) -> vector<16xindex> {
   %0 = vector.flat_transpose %arg0 { rows = 4: i32, columns = 4: i32 }
      : vector<16xindex> -> vector<16xindex>
   return %0 : vector<16xindex>
 }
 // CHECK-LABEL: func @flat_transpose_index
 // CHECK-SAME:  %[[A:.*]]: vector<16xindex>
-// CHECK:       %[[T0:.*]] = unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
+// CHECK:       %[[T0:.*]] = builtin.unrealized_conversion_cast %[[A]] : vector<16xindex> to vector<16xi64>
 // CHECK:       %[[T1:.*]] = llvm.intr.matrix.transpose %[[T0]]
 // CHECK-SAME:      {columns = 4 : i32, rows = 4 : i32} :
 // CHECK-SAME:      vector<16xi64> into vector<16xi64>
-// CHECK:       %[[T2:.*]] = unrealized_conversion_cast %[[T1]] : vector<16xi64> to vector<16xindex>
+// CHECK:       %[[T2:.*]] = builtin.unrealized_conversion_cast %[[T1]] : vector<16xi64> to vector<16xindex>
 // CHECK:       return %[[T2]] : vector<16xindex>
 
 // -----
 
 func @vector_load_op(%memref : memref<200x100xf32>, %i : index, %j : index) -> vector<8xf32> {
   %0 = vector.load %memref[%i, %j] : memref<200x100xf32>, vector<8xf32>
   return %0 : vector<8xf32>
 }

 // -----
 
 func @vector_load_op_index(%memref : memref<200x100xindex>, %i : index, %j : index) -> vector<8xindex> {
   %0 = vector.load %memref[%i, %j] : memref<200x100xindex>, vector<8xindex>
   return %0 : vector<8xindex>
 }
 // CHECK-LABEL: func @vector_load_op_index
 // CHECK: %[[T0:.*]] = llvm.load %{{.*}} {alignment = 8 : i64} : !llvm.ptr<vector<8xi64>>
-// CHECK: %[[T1:.*]] = unrealized_conversion_cast %[[T0]] : vector<8xi64> to vector<8xindex>
+// CHECK: %[[T1:.*]] = builtin.unrealized_conversion_cast %[[T0]] : vector<8xi64> to vector<8xindex>
 // CHECK: return %[[T1]] : vector<8xindex>
 
 // -----
 
 func @vector_load_op_aligned(%memref : memref<200x100xf32>, %i : index, %j : index) -> vector<8xf32> {
   memref.assume_alignment %memref, 32 : memref<200x100xf32>
   %0 = vector.load %memref[%i, %j] : memref<200x100xf32>, vector<8xf32>
   return %0 : vector<8xf32>

 func @masked_load_op(%arg0: memref<?xf32>, %arg1: vector<16xi1>, %arg2: vector<16xf32>) -> vector<16xf32> {
   %c0 = constant 0: index
   %0 = vector.maskedload %arg0[%c0], %arg1, %arg2 : memref<?xf32>, vector<16xi1>, vector<16xf32> into vector<16xf32>
   return %0 : vector<16xf32>
 }
 
 // CHECK-LABEL: func @masked_load_op
 // CHECK: %[[CO:.*]] = constant 0 : index
-// CHECK: %[[C:.*]] = unrealized_conversion_cast %[[CO]] : index to i64
+// CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[CO]] : index to i64
 // CHECK: %[[P:.*]] = llvm.getelementptr %{{.*}}[%[[C]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK: %[[B:.*]] = llvm.bitcast %[[P]] : !llvm.ptr<f32> to !llvm.ptr<vector<16xf32>>
 // CHECK: %[[L:.*]] = llvm.intr.masked.load %[[B]], %{{.*}}, %{{.*}} {alignment = 4 : i32} : (!llvm.ptr<vector<16xf32>>, vector<16xi1>, vector<16xf32>) -> vector<16xf32>
 // CHECK: return %[[L]] : vector<16xf32>
 
 // -----
 
 func @masked_load_op_index(%arg0: memref<?xindex>, %arg1: vector<16xi1>, %arg2: vector<16xindex>) -> vector<16xindex> {

 func @masked_store_op(%arg0: memref<?xf32>, %arg1: vector<16xi1>, %arg2: vector<16xf32>) {
   %c0 = constant 0: index
   vector.maskedstore %arg0[%c0], %arg1, %arg2 : memref<?xf32>, vector<16xi1>, vector<16xf32>
   return
 }
 
 // CHECK-LABEL: func @masked_store_op
 // CHECK: %[[CO:.*]] = constant 0 : index
-// CHECK: %[[C:.*]] = unrealized_conversion_cast %[[CO]] : index to i64
+// CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[CO]] : index to i64
 // CHECK: %[[P:.*]] = llvm.getelementptr %{{.*}}[%[[C]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK: %[[B:.*]] = llvm.bitcast %[[P]] : !llvm.ptr<f32> to !llvm.ptr<vector<16xf32>>
 // CHECK: llvm.intr.masked.store %{{.*}}, %[[B]], %{{.*}} {alignment = 4 : i32} : vector<16xf32>, vector<16xi1> into !llvm.ptr<vector<16xf32>>
 
 // -----
 
 func @masked_store_op_index(%arg0: memref<?xindex>, %arg1: vector<16xi1>, %arg2: vector<16xindex>) {
   %c0 = constant 0: index
func @gather_op_index(%arg0: memref<?xindex>, %arg1: vector<3xindex>, %arg2: vector<3xi1>, %arg3: vector<3xindex>) -> vector<3xindex> {
   %0 = constant 0: index
   %1 = vector.gather %arg0[%0][%arg1], %arg2, %arg3 : memref<?xindex>, vector<3xindex>, vector<3xi1>, vector<3xindex> into vector<3xindex>
   return %1 : vector<3xindex>
 }
 
 // CHECK-LABEL: func @gather_op_index
 // CHECK: %[[P:.*]] = llvm.getelementptr %{{.*}}[%{{.*}}] : (!llvm.ptr<i64>, vector<3xi64>) -> !llvm.vec<3 x ptr<i64>>
 // CHECK: %[[G:.*]] = llvm.intr.masked.gather %{{.*}}, %{{.*}}, %{{.*}} {alignment = 8 : i32} : (!llvm.vec<3 x ptr<i64>>, vector<3xi1>, vector<3xi64>) -> vector<3xi64>
-// CHECK: %{{.*}} = unrealized_conversion_cast %[[G]] : vector<3xi64> to vector<3xindex>
+// CHECK: %{{.*}} = builtin.unrealized_conversion_cast %[[G]] : vector<3xi64> to vector<3xindex>
 
 // -----
 
 func @gather_op_aligned(%arg0: memref<?xf32>, %arg1: vector<3xi32>, %arg2: vector<3xi1>, %arg3: vector<3xf32>) -> vector<3xf32> {
   memref.assume_alignment %arg0, 32 : memref<?xf32>
   %0 = constant 0: index
   %1 = vector.gather %arg0[%0][%arg1], %arg2, %arg3 : memref<?xf32>, vector<3xi32>, vector<3xi1>, vector<3xf32> into vector<3xf32>
   return %1 : vector<3xf32>

 func @expand_load_op(%arg0: memref<?xf32>, %arg1: vector<11xi1>, %arg2: vector<11xf32>) -> vector<11xf32> {
   %c0 = constant 0: index
   %0 = vector.expandload %arg0[%c0], %arg1, %arg2 : memref<?xf32>, vector<11xi1>, vector<11xf32> into vector<11xf32>
   return %0 : vector<11xf32>
 }
 
 // CHECK-LABEL: func @expand_load_op
 // CHECK: %[[CO:.*]] = constant 0 : index
-// CHECK: %[[C:.*]] = unrealized_conversion_cast %[[CO]] : index to i64
+// CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[CO]] : index to i64
 // CHECK: %[[P:.*]] = llvm.getelementptr %{{.*}}[%[[C]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK: %[[E:.*]] = "llvm.intr.masked.expandload"(%[[P]], %{{.*}}, %{{.*}}) : (!llvm.ptr<f32>, vector<11xi1>, vector<11xf32>) -> vector<11xf32>
 // CHECK: return %[[E]] : vector<11xf32>
 
 // -----
 
 func @expand_load_op_index(%arg0: memref<?xindex>, %arg1: vector<11xi1>, %arg2: vector<11xindex>) -> vector<11xindex> {
   %c0 = constant 0: index
   %0 = vector.expandload %arg0[%c0], %arg1, %arg2 : memref<?xindex>, vector<11xi1>, vector<11xindex> into vector<11xindex>
   return %0 : vector<11xindex>
 }
 // CHECK-LABEL: func @expand_load_op_index
 // CHECK: %{{.*}} = "llvm.intr.masked.expandload"(%{{.*}}, %{{.*}}, %{{.*}}) : (!llvm.ptr<i64>, vector<11xi1>, vector<11xi64>) -> vector<11xi64>
 
 // -----
 
 func @compress_store_op(%arg0: memref<?xf32>, %arg1: vector<11xi1>, %arg2: vector<11xf32>) {
   %c0 = constant 0: index
   vector.compressstore %arg0[%c0], %arg1, %arg2 : memref<?xf32>, vector<11xi1>, vector<11xf32>
   return
 }
 
 // CHECK-LABEL: func @compress_store_op
 // CHECK: %[[CO:.*]] = constant 0 : index
-// CHECK: %[[C:.*]] = unrealized_conversion_cast %[[CO]] : index to i64
+// CHECK: %[[C:.*]] = builtin.unrealized_conversion_cast %[[CO]] : index to i64
 // CHECK: %[[P:.*]] = llvm.getelementptr %{{.*}}[%[[C]]] : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
 // CHECK: "llvm.intr.masked.compressstore"(%{{.*}}, %[[P]], %{{.*}}) : (vector<11xf32>, !llvm.ptr<f32>, vector<11xi1>) -> ()
 
 // -----
 
 func @compress_store_op_index(%arg0: memref<?xindex>, %arg1: vector<11xi1>, %arg2: vector<11xindex>) {
   %c0 = constant 0: index
   vector.compressstore %arg0[%c0], %arg1, %arg2 : memref<?xindex>, vector<11xi1>, vector<11xindex>
   return
 }
 // CHECK-LABEL: func @compress_store_op_index
 // CHECK: "llvm.intr.masked.compressstore"(%{{.*}}, %{{.*}}, %{{.*}}) : (vector<11xi64>, !llvm.ptr<i64>, vector<11xi1>) -> ()

mlir/test/Dialect/Affine/canonicalize.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -split-input-file -pass-pipeline='func(canonicalize)' | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -split-input-file -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s
 
 // -----
 
 // CHECK-DAG: #[[$MAP0:.*]] = affine_map<(d0) -> (d0 - 1)>
 // CHECK-DAG: #[[$MAP1:.*]] = affine_map<(d0) -> (d0 + 1)>
 
 // CHECK-LABEL: func @compose_affine_maps_1dto2d_no_symbols() {
 func @compose_affine_maps_1dto2d_no_symbols() {

mlir/test/Dialect/ArmSVE/memcpy.mlir

-// RUN: mlir-opt %s -convert-vector-to-llvm="enable-arm-sve" | mlir-opt | FileCheck %s 
+// RUN: mlir-opt %s -convert-vector-to-llvm="enable-arm-sve" | mlir-opt | FileCheck %s
 
 // CHECK: memcopy([[SRC:%arg[0-9]+]]: memref<?xf32>, [[DST:%arg[0-9]+]]
 func @memcopy(%src : memref<?xf32>, %dst : memref<?xf32>, %size : index) {
   %c0 = constant 0 : index
   %c4 = constant 4 : index
   %vs = arm_sve.vector_scale : index
   %step = muli %c4, %vs : index
 
-  // CHECK: scf.for [[LOOPIDX:%arg[0-9]+]] = {{.*}} 
+  // CHECK: scf.for [[LOOPIDX:%arg[0-9]+]] = {{.*}}
   scf.for %i0 = %c0 to %size step %step {
-    // CHECK: [[SRCMRS:%[0-9]+]] = unrealized_conversion_cast [[SRC]] : memref<?xf32> to !llvm.struct<(ptr<f32>
-    // CHECK: [[SRCIDX:%[0-9]+]] = unrealized_conversion_cast [[LOOPIDX]] : index to i64
+    // CHECK: [[SRCMRS:%[0-9]+]] = builtin.unrealized_conversion_cast [[SRC]] : memref<?xf32> to !llvm.struct<(ptr<f32>
+    // CHECK: [[SRCIDX:%[0-9]+]] = builtin.unrealized_conversion_cast [[LOOPIDX]] : index to i64
     // CHECK: [[SRCMEM:%[0-9]+]] = llvm.extractvalue [[SRCMRS]][1] : !llvm.struct<(ptr<f32>
     // CHECK-NEXT: [[SRCPTR:%[0-9]+]] = llvm.getelementptr [[SRCMEM]]{{.}}[[SRCIDX]]{{.}} : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
     // CHECK-NEXT: [[SRCVPTR:%[0-9]+]] = llvm.bitcast [[SRCPTR]] : !llvm.ptr<f32> to !llvm.ptr<vec<? x 4 x f32>>
     // CHECK-NEXT: [[LDVAL:%[0-9]+]] = llvm.load [[SRCVPTR]] : !llvm.ptr<vec<? x 4 x f32>>
     %0 = arm_sve.load %src[%i0] : !arm_sve.vector<4xf32> from memref<?xf32>
-    // CHECK: [[DSTMRS:%[0-9]+]] = unrealized_conversion_cast [[DST]] : memref<?xf32> to !llvm.struct<(ptr<f32>
-    // CHECK: [[DSTIDX:%[0-9]+]] = unrealized_conversion_cast [[LOOPIDX]] : index to i64
+    // CHECK: [[DSTMRS:%[0-9]+]] = builtin.unrealized_conversion_cast [[DST]] : memref<?xf32> to !llvm.struct<(ptr<f32>
+    // CHECK: [[DSTIDX:%[0-9]+]] = builtin.unrealized_conversion_cast [[LOOPIDX]] : index to i64
     // CHECK: [[DSTMEM:%[0-9]+]] = llvm.extractvalue [[DSTMRS]][1] : !llvm.struct<(ptr<f32>
     // CHECK-NEXT: [[DSTPTR:%[0-9]+]] = llvm.getelementptr [[DSTMEM]]{{.}}[[DSTIDX]]{{.}} : (!llvm.ptr<f32>, i64) -> !llvm.ptr<f32>
     // CHECK-NEXT: [[DSTVPTR:%[0-9]+]] = llvm.bitcast [[DSTPTR]] : !llvm.ptr<f32> to !llvm.ptr<vec<? x 4 x f32>>
     // CHECK-NEXT: llvm.store [[LDVAL]], [[DSTVPTR]] : !llvm.ptr<vec<? x 4 x f32>>
     arm_sve.store %0, %dst[%i0] : !arm_sve.vector<4xf32> to memref<?xf32>
   }
 
   return
 }

mlir/test/Dialect/Builtin/invalid.mlir

 // RUN: mlir-opt %s -split-input-file -verify-diagnostics
 
 //===----------------------------------------------------------------------===//
 // UnrealizedConversionCastOp
 //===----------------------------------------------------------------------===//
 
 // expected-error@+1 {{expected at least one result for cast operation}}
-"unrealized_conversion_cast"() : () -> ()
+"builtin.unrealized_conversion_cast"() : () -> ()
 
 // -----
 

mlir/test/Dialect/LLVMIR/terminator.mlir

-// RUN: mlir-opt -pass-pipeline='func(canonicalize)' %s | FileCheck %s
+// RUN: mlir-opt -pass-pipeline='builtin.func(canonicalize)' %s | FileCheck %s
 // verify that terminators survive the canonicalizer
 
 // CHECK-LABEL: @return
 // CHECK: llvm.return
 func @return() {
   llvm.return
 }
 

mlir/test/Dialect/Linalg/fusion-sequence.mlir

-// RUN: mlir-opt -pass-pipeline="func(test-linalg-tile-and-fuse{tile-sizes=16,32,64}),resolve-shaped-type-result-dims,canonicalize,cse" -split-input-file %s | FileCheck %s
+// RUN: mlir-opt -pass-pipeline="builtin.func(test-linalg-tile-and-fuse{tile-sizes=16,32,64}),resolve-shaped-type-result-dims,canonicalize,cse" -split-input-file %s | FileCheck %s
 
 module {
   func @three_op_fusion(%arg0: memref<?x?xf32>, %arg1: memref<?x?xf32>,
                         %arg2: memref<?xf32>, %arg3 : memref<?x?xf32>) {
     %cst = constant 0.000000e+00 : f32
     %c0 = constant 0 : index
     %c1 = constant 1 : index
     %d0 = memref.dim %arg0, %c0 : memref<?x?xf32>

mlir/test/Dialect/Quant/canonicalize.mlir

-// RUN: mlir-opt %s -split-input-file -pass-pipeline='func(canonicalize)' | FileCheck %s
+// RUN: mlir-opt %s -split-input-file -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s
 
 // -----
 // CHECK-LABEL: redundant_scast
 func @redundant_scast() -> tensor<4xi8> {
   // CHECK-NEXT: constant dense<10> : tensor<4xi8>
   // CHECK-NEXT: return
   %cst = constant dense<5> : tensor<4xi8>
   %1 = "quant.scast"(%cst) : (tensor<4xi8>) -> tensor<4x!quant.uniform<u8:f32, 7.812500e-03:128>>

mlir/test/Dialect/SCF/canonicalize.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize)' -split-input-file | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize)' -split-input-file | FileCheck %s
 
 
 // -----
 
 func @single_iteration_some(%A: memref<?x?x?xi32>) {
   %c0 = constant 0 : index
   %c1 = constant 1 : index
   %c2 = constant 2 : index

mlir/test/Dialect/SCF/loop-range.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(for-loop-range-folding)' -split-input-file | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(for-loop-range-folding)' -split-input-file | FileCheck %s
 
 func @fold_one_loop(%arg0: memref<?xi32>, %arg1: index, %arg2: index) {
   %c0 = constant 0 : index
   %c1 = constant 1 : index
   %c4 = constant 4 : index
   scf.for %i = %c0 to %arg1 step %c1 {
     %0 = addi %arg2, %i : index
     %1 = muli %0, %c4 : index

mlir/test/Dialect/SCF/parallel-loop-fusion.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(parallel-loop-fusion)' -split-input-file | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(parallel-loop-fusion)' -split-input-file | FileCheck %s
 
 func @fuse_empty_loops() {
   %c2 = constant 2 : index
   %c0 = constant 0 : index
   %c1 = constant 1 : index
   scf.parallel (%i, %j) = (%c0, %c0) to (%c2, %c2) step (%c1, %c1) {
     scf.yield
   }

mlir/test/Dialect/SCF/parallel-loop-tiling.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(parallel-loop-tiling{parallel-loop-tile-sizes=1,4})' -split-input-file | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(parallel-loop-tiling{parallel-loop-tile-sizes=1,4})' -split-input-file | FileCheck %s
 
 func @parallel_loop(%arg0 : index, %arg1 : index, %arg2 : index,
                     %arg3 : index, %arg4 : index, %arg5 : index,
         %A: memref<?x?xf32>, %B: memref<?x?xf32>,
                     %C: memref<?x?xf32>, %result: memref<?x?xf32>) {
   scf.parallel (%i0, %i1) = (%arg0, %arg1) to (%arg2, %arg3) step (%arg4, %arg5) {
     %B_elem = memref.load %B[%i0, %i1] : memref<?x?xf32>
     %C_elem = memref.load %C[%i0, %i1] : memref<?x?xf32>

mlir/test/Dialect/SPIRV/Transforms/canonicalize.mlir

-// RUN: mlir-opt %s -split-input-file -pass-pipeline='func(canonicalize)' | FileCheck %s
+// RUN: mlir-opt %s -split-input-file -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s
 
 //===----------------------------------------------------------------------===//
 // spv.AccessChain
 //===----------------------------------------------------------------------===//
 
 func @combine_full_access_chain() -> f32 {
   // CHECK: %[[INDEX:.*]] = spv.Constant 0
   // CHECK-NEXT: %[[VAR:.*]] = spv.Variable

mlir/test/Dialect/Standard/canonicalize-cf.mlir

-// RUN: mlir-opt %s -allow-unregistered-dialect -pass-pipeline='func(canonicalize)' -split-input-file | FileCheck --dump-input-context 20 %s
+// RUN: mlir-opt %s -allow-unregistered-dialect -pass-pipeline='builtin.func(canonicalize)' -split-input-file | FileCheck --dump-input-context 20 %s
 
 /// Test the folding of BranchOp.
 
 // CHECK-LABEL: func @br_folding(
 func @br_folding() -> i32 {
   // CHECK-NEXT: %[[CST:.*]] = constant 0 : i32
   // CHECK-NEXT: return %[[CST]] : i32
   %c0_i32 = constant 0 : i32

mlir/test/Dialect/Vector/canonicalize.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize)' -split-input-file -allow-unregistered-dialect | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize)' -split-input-file -allow-unregistered-dialect | FileCheck %s
 
 // -----
 
 // CHECK-LABEL: create_vector_mask_to_constant_mask
 func @create_vector_mask_to_constant_mask() -> (vector<4x3xi1>) {
   %c2 = constant 2 : index
   %c3 = constant 3 : index
   // CHECK: vector.constant_mask [3, 2] : vector<4x3xi1>

mlir/test/IR/invalid-ops.mlir

func @trunci_cast_to_same_width(%arg0 : i16) {
   %0 = trunci %arg0 : i16 to i16
   return
 }
 
 // -----
 
 func @return_not_in_function() {
   "foo.region"() ({
-    // expected-error@+1 {{'std.return' op expects parent op 'func'}}
+    // expected-error@+1 {{'std.return' op expects parent op 'builtin.func'}}
     return
   }): () -> ()
   return
 }
 
 // -----
 
 func @invalid_splat(%v : f32) {

mlir/test/IR/print-ir-defuse.mlir

 // CHECK:   - Operand produced by operation 'dialect.op1'
 // CHECK: Has 0 results:
 // CHECK: Visiting op 'dialect.innerop1' with 2 operands:
 // CHECK:   - Operand produced by Block argument, number 0
 // CHECK:   - Operand produced by operation 'dialect.op1'
 // CHECK: Has 0 results:
 // CHECK: Visiting op 'dialect.op3' with 0 operands:
 // CHECK: Has 0 results:
-// CHECK: Visiting op 'module' with 0 operands:
+// CHECK: Visiting op 'builtin.module' with 0 operands:
 // CHECK: Has 0 results:
 
 %results:4 = "dialect.op1"() : () -> (i1, i16, i32, i64)
 "dialect.op2"(%results#0, %results#2) : (i1, i32) -> ()
 "dialect.op3"() ({
   ^bb0(%arg0 : i1):
     "dialect.innerop1"(%arg0, %results#2) : (i1, i32) -> ()
 }) : () -> ()

mlir/test/IR/print-ir-nesting.mlir

 // RUN: mlir-opt -test-print-nesting  -allow-unregistered-dialect %s | FileCheck %s
 
-// CHECK: visiting op: 'module' with 0 operands and 0 results
+// CHECK: visiting op: 'builtin.module' with 0 operands and 0 results
 // CHECK:  1 nested regions:
 // CHECK:   Region with 1 blocks:
 // CHECK:     Block with 0 arguments, 0 successors, and 2 operations
 module {
 
 
 // CHECK:       visiting op: 'dialect.op1' with 0 operands and 4 results
 // CHECK:       1 attributes:

mlir/test/IR/print-op-on-diagnostic.mlir

 // RUN: not mlir-opt %s -mlir-print-op-on-diagnostic 2>&1 | FileCheck %s
 
 // This file tests the functionality of 'mlir-print-op-on-diagnostic'.
 
 // CHECK: {{invalid to use 'test.invalid_attr'}}
-// CHECK: {{see current operation: "module"()}}
+// CHECK: {{see current operation: "builtin.module"()}}
 module attributes {test.invalid_attr} {}

mlir/test/IR/visitors.mlir

scf.if %cond {
       "use2"(%i) : (index) -> ()
     }
     "use3"(%i) : (index) -> ()
   }
   return
 }
 
 // CHECK-LABEL: Op pre-order visit
-// CHECK:       Visiting op 'module'
-// CHECK:       Visiting op 'func'
+// CHECK:       Visiting op 'builtin.module'
+// CHECK:       Visiting op 'builtin.func'
 // CHECK:       Visiting op 'scf.for'
 // CHECK:       Visiting op 'use0'
 // CHECK:       Visiting op 'scf.if'
 // CHECK:       Visiting op 'use1'
 // CHECK:       Visiting op 'use2'
 // CHECK:       Visiting op 'use3'
 // CHECK:       Visiting op 'std.return'
 
 // CHECK-LABEL: Block pre-order visits
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'module'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'func'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.module'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.func'
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'scf.for'
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'scf.if'
 // CHECK:       Visiting block ^bb0 from region 1 from operation 'scf.if'
 
 // CHECK-LABEL: Region pre-order visits
-// CHECK:       Visiting region 0 from operation 'module'
-// CHECK:       Visiting region 0 from operation 'func'
+// CHECK:       Visiting region 0 from operation 'builtin.module'
+// CHECK:       Visiting region 0 from operation 'builtin.func'
 // CHECK:       Visiting region 0 from operation 'scf.for'
 // CHECK:       Visiting region 0 from operation 'scf.if'
 // CHECK:       Visiting region 1 from operation 'scf.if'
 
 // CHECK-LABEL: Op post-order visits
 // CHECK:       Visiting op 'use0'
 // CHECK:       Visiting op 'use1'
 // CHECK:       Visiting op 'use2'
 // CHECK:       Visiting op 'scf.if'
 // CHECK:       Visiting op 'use3'
 // CHECK:       Visiting op 'scf.for'
 // CHECK:       Visiting op 'std.return'
-// CHECK:       Visiting op 'func'
-// CHECK:       Visiting op 'module'
+// CHECK:       Visiting op 'builtin.func'
+// CHECK:       Visiting op 'builtin.module'
 
 // CHECK-LABEL: Block post-order visits
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'scf.if'
 // CHECK:       Visiting block ^bb0 from region 1 from operation 'scf.if'
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'scf.for'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'func'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'module'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.func'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.module'
 
 // CHECK-LABEL: Region post-order visits
 // CHECK:       Visiting region 0 from operation 'scf.if'
 // CHECK:       Visiting region 1 from operation 'scf.if'
 // CHECK:       Visiting region 0 from operation 'scf.for'
-// CHECK:       Visiting region 0 from operation 'func'
-// CHECK:       Visiting region 0 from operation 'module'
+// CHECK:       Visiting region 0 from operation 'builtin.func'
+// CHECK:       Visiting region 0 from operation 'builtin.module'
 
 // CHECK-LABEL: Op pre-order erasures
 // CHECK:       Erasing op 'scf.for'
 // CHECK:       Erasing op 'std.return'
 
 // CHECK-LABEL: Block pre-order erasures
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'scf.for'
 

 // CHECK-LABEL: Op post-order erasures (no skip)
 // CHECK:       Erasing op 'use0'
 // CHECK:       Erasing op 'use1'
 // CHECK:       Erasing op 'use2'
 // CHECK:       Erasing op 'scf.if'
 // CHECK:       Erasing op 'use3'
 // CHECK:       Erasing op 'scf.for'
 // CHECK:       Erasing op 'std.return'
-// CHECK:       Erasing op 'func'
-// CHECK:       Erasing op 'module'
+// CHECK:       Erasing op 'builtin.func'
+// CHECK:       Erasing op 'builtin.module'
 
 // CHECK-LABEL: Block post-order erasures (no skip)
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'scf.if'
 // CHECK:       Erasing block ^bb0 from region 1 from operation 'scf.if'
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'scf.for'
-// CHECK:       Erasing block ^bb0 from region 0 from operation 'func'
-// CHECK:       Erasing block ^bb0 from region 0 from operation 'module'
+// CHECK:       Erasing block ^bb0 from region 0 from operation 'builtin.func'
+// CHECK:       Erasing block ^bb0 from region 0 from operation 'builtin.module'
 
 // -----
 
 func @unstructured_cfg() {
   "regionOp0"() ({
     ^bb0:
       "op0"() : () -> ()
       br ^bb2
     ^bb1:
       "op1"() : () -> ()
       br ^bb2
     ^bb2:
       "op2"() : () -> ()
   }) : () -> ()
   return
 }
 
 // CHECK-LABEL: Op pre-order visits
-// CHECK:       Visiting op 'module'
-// CHECK:       Visiting op 'func'
+// CHECK:       Visiting op 'builtin.module'
+// CHECK:       Visiting op 'builtin.func'
 // CHECK:       Visiting op 'regionOp0'
 // CHECK:       Visiting op 'op0'
 // CHECK:       Visiting op 'std.br'
 // CHECK:       Visiting op 'op1'
 // CHECK:       Visiting op 'std.br'
 // CHECK:       Visiting op 'op2'
 // CHECK:       Visiting op 'std.return'
 
 // CHECK-LABEL: Block pre-order visits
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'module'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'func'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.module'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.func'
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'regionOp0'
 // CHECK:       Visiting block ^bb1 from region 0 from operation 'regionOp0'
 // CHECK:       Visiting block ^bb2 from region 0 from operation 'regionOp0'
 
 // CHECK-LABEL: Region pre-order visits
-// CHECK:       Visiting region 0 from operation 'module'
-// CHECK:       Visiting region 0 from operation 'func'
+// CHECK:       Visiting region 0 from operation 'builtin.module'
+// CHECK:       Visiting region 0 from operation 'builtin.func'
 // CHECK:       Visiting region 0 from operation 'regionOp0'
 
 // CHECK-LABEL: Op post-order visits
 // CHECK:       Visiting op 'op0'
 // CHECK:       Visiting op 'std.br'
 // CHECK:       Visiting op 'op1'
 // CHECK:       Visiting op 'std.br'
 // CHECK:       Visiting op 'op2'
 // CHECK:       Visiting op 'regionOp0'
 // CHECK:       Visiting op 'std.return'
-// CHECK:       Visiting op 'func'
-// CHECK:       Visiting op 'module'
+// CHECK:       Visiting op 'builtin.func'
+// CHECK:       Visiting op 'builtin.module'
 
 // CHECK-LABEL: Block post-order visits
 // CHECK:       Visiting block ^bb0 from region 0 from operation 'regionOp0'
 // CHECK:       Visiting block ^bb1 from region 0 from operation 'regionOp0'
 // CHECK:       Visiting block ^bb2 from region 0 from operation 'regionOp0'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'func'
-// CHECK:       Visiting block ^bb0 from region 0 from operation 'module'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.func'
+// CHECK:       Visiting block ^bb0 from region 0 from operation 'builtin.module'
 
 // CHECK-LABEL: Region post-order visits
 // CHECK:       Visiting region 0 from operation 'regionOp0'
-// CHECK:       Visiting region 0 from operation 'func'
-// CHECK:       Visiting region 0 from operation 'module'
+// CHECK:       Visiting region 0 from operation 'builtin.func'
+// CHECK:       Visiting region 0 from operation 'builtin.module'
 
 // CHECK-LABEL: Op pre-order erasures (skip)
 // CHECK:       Erasing op 'regionOp0'
 // CHECK:       Erasing op 'std.return'
 
 // CHECK-LABEL: Block pre-order erasures (skip)
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'regionOp0'
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'regionOp0'

 // CHECK:       Erasing op 'op2'
 // CHECK:       Erasing op 'regionOp0'
 // CHECK:       Erasing op 'std.return'
 
 // CHECK-LABEL: Block post-order erasures (no skip)
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'regionOp0'
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'regionOp0'
 // CHECK:       Erasing block ^bb0 from region 0 from operation 'regionOp0'
-// CHECK:       Erasing block ^bb0 from region 0 from operation 'func'
-// CHECK:       Erasing block ^bb0 from region 0 from operation 'module'
+// CHECK:       Erasing block ^bb0 from region 0 from operation 'builtin.func'
+// CHECK:       Erasing block ^bb0 from region 0 from operation 'builtin.module'

mlir/test/IR/wrapping_op.mlir

 // RUN: mlir-opt -allow-unregistered-dialect %s | FileCheck %s
 // RUN: mlir-opt -allow-unregistered-dialect -mlir-print-op-generic -mlir-print-debuginfo -mlir-print-local-scope %s | FileCheck %s --check-prefix=CHECK-GENERIC
 
 // CHECK-LABEL: func @wrapping_op
-// CHECK-GENERIC: "func"
+// CHECK-GENERIC: "builtin.func"
 func @wrapping_op(%arg0 : i32, %arg1 : f32) -> (i3, i2, i1) {
 // CHECK: %0:3 = test.wrapping_region wraps "some.op"(%arg1, %arg0) {test.attr = "attr"} : (f32, i32) -> (i1, i2, i3)
 // CHECK-GENERIC: "test.wrapping_region"() ( {
 // CHECK-GENERIC:   %[[NESTED_RES:.*]]:3 = "some.op"(%arg1, %arg0) {test.attr = "attr"} : (f32, i32) -> (i1, i2, i3) loc("some_NameLoc")
 // CHECK-GENERIC:   "test.return"(%[[NESTED_RES]]#0, %[[NESTED_RES]]#1, %[[NESTED_RES]]#2) : (i1, i2, i3) -> () loc("some_NameLoc")
 // CHECK-GENERIC: }) : () -> (i1, i2, i3) loc("some_NameLoc")
   %res:3 = test.wrapping_region wraps "some.op"(%arg1, %arg0) { test.attr = "attr" } : (f32, i32) -> (i1, i2, i3) loc("some_NameLoc")
   return %res#2, %res#1, %res#0 : i3, i2, i1
 }
 // CHECK-GENERIC: sym_name = "wrapping_op"

mlir/test/Pass/crash-recovery.mlir

-// RUN: mlir-opt %s -pass-pipeline='module(test-module-pass, test-pass-crash)' -pass-pipeline-crash-reproducer=%t -verify-diagnostics
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-module-pass, test-pass-crash)' -pass-pipeline-crash-reproducer=%t -verify-diagnostics
 // RUN: cat %t | FileCheck -check-prefix=REPRO %s
-// RUN: mlir-opt %s -pass-pipeline='module(test-module-pass, test-pass-crash)' -pass-pipeline-crash-reproducer=%t -verify-diagnostics -pass-pipeline-local-reproducer -mlir-disable-threading
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-module-pass, test-pass-crash)' -pass-pipeline-crash-reproducer=%t -verify-diagnostics -pass-pipeline-local-reproducer -mlir-disable-threading
 // RUN: cat %t | FileCheck -check-prefix=REPRO_LOCAL %s
 
 // Check that we correctly handle verifiers passes with local reproducer, this used to crash.
 // RUN: mlir-opt %s -test-module-pass -test-module-pass  -test-module-pass -pass-pipeline-crash-reproducer=%t -pass-pipeline-local-reproducer -mlir-disable-threading
 // RUN: cat %t | FileCheck -check-prefix=REPRO_LOCAL %s
 
 // Check that local reproducers will also traverse dynamic pass pipelines.
 // RUN: mlir-opt %s -pass-pipeline='test-module-pass,test-dynamic-pipeline{op-name=inner_mod1 run-on-nested-operations=1 dynamic-pipeline=test-pass-crash}' -pass-pipeline-crash-reproducer=%t -verify-diagnostics -pass-pipeline-local-reproducer --mlir-disable-threading
 // RUN: cat %t | FileCheck -check-prefix=REPRO_LOCAL_DYNAMIC %s
 
 // expected-error@below {{Failures have been detected while processing an MLIR pass pipeline}}
 // expected-note@below {{Pipeline failed while executing}}
 module @inner_mod1 {
   module @foo {}
 }
 
-// REPRO: configuration: -pass-pipeline='module(test-module-pass, test-pass-crash)'
+// REPRO: configuration: -pass-pipeline='builtin.module(test-module-pass, test-pass-crash)'
 
 // REPRO: module @inner_mod1
 // REPRO: module @foo {
 
-// REPRO_LOCAL: configuration: -pass-pipeline='module(test-pass-crash)'
+// REPRO_LOCAL: configuration: -pass-pipeline='builtin.module(test-pass-crash)'
 
 // REPRO_LOCAL: module @inner_mod1
 // REPRO_LOCAL: module @foo {
 
-// REPRO_LOCAL_DYNAMIC: configuration: -pass-pipeline='module(test-pass-crash)'
+// REPRO_LOCAL_DYNAMIC: configuration: -pass-pipeline='builtin.module(test-pass-crash)'
 
 // REPRO_LOCAL_DYNAMIC: module @inner_mod1
 // REPRO_LOCAL_DYNAMIC: module @foo {

mlir/test/Pass/dynamic-pipeline-fail-on-parent.mlir

-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod1 run-on-parent=1 dynamic-pipeline=test-patterns})'  -split-input-file -verify-diagnostics
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod1 run-on-parent=1 dynamic-pipeline=test-patterns})'  -split-input-file -verify-diagnostics
 
 // Verify that we fail to schedule a dynamic pipeline on the parent operation.
 
-// expected-error @+1 {{'module' op Trying to schedule a dynamic pipeline on an operation that isn't nested under the current operation}}
+// expected-error @+1 {{'builtin.module' op Trying to schedule a dynamic pipeline on an operation that isn't nested under the current operation}}
 module {
 module @inner_mod1 {
   "test.symbol"() {sym_name = "foo"} : () -> ()
   func private @bar()
 }
 }

mlir/test/Pass/dynamic-pipeline-nested.mlir

-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod1 dynamic-pipeline=cse})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=NOTNESTED --check-prefix=CHECK
-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod1 run-on-nested-operations=1 dynamic-pipeline=cse})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=NESTED --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod1 dynamic-pipeline=cse})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=NOTNESTED --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod1 run-on-nested-operations=1 dynamic-pipeline=cse})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=NESTED --check-prefix=CHECK
 
 
 // Verify that we can schedule a dynamic pipeline on a nested operation
 
 func @f() {
   return
 }
 

mlir/test/Pass/dynamic-pipeline.mlir

-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod1, dynamic-pipeline=func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD1-ONLY --check-prefix=CHECK
-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod2, dynamic-pipeline=func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD2 --check-prefix=MOD2-ONLY --check-prefix=CHECK
-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{op-name=inner_mod1,inner_mod2, dynamic-pipeline=func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD2 --check-prefix=CHECK
-// RUN: mlir-opt %s -pass-pipeline='module(test-dynamic-pipeline{dynamic-pipeline=func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD2 --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod1, dynamic-pipeline=builtin.func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD1-ONLY --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod2, dynamic-pipeline=builtin.func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD2 --check-prefix=MOD2-ONLY --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{op-name=inner_mod1,inner_mod2, dynamic-pipeline=builtin.func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD2 --check-prefix=CHECK
+// RUN: mlir-opt %s -pass-pipeline='builtin.module(test-dynamic-pipeline{dynamic-pipeline=builtin.func(cse,canonicalize)})'  --mlir-disable-threading  -print-ir-before-all 2>&1 | FileCheck %s --check-prefix=MOD1 --check-prefix=MOD2 --check-prefix=CHECK
 
 
 func @f() {
   return
 }
 
 // CHECK: IR Dump Before
 // CHECK-SAME: TestDynamicPipelinePass

mlir/test/Pass/invalid-pass.mlir

-// RUN: not mlir-opt %s -pass-pipeline='module(test-module-pass{test-option=a})' 2>&1 | FileCheck %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(test-module-pass{test-option=a})' 2>&1 | FileCheck %s
 
 // CHECK: <Pass-Options-Parser>: no such option test-option
 // CHECK: failed to add `test-module-pass` with options `test-option=a`
-// CHECK: failed to add `module` with options `` to inner pipeline
+// CHECK: failed to add `builtin.module` with options `` to inner pipeline
 module {}

mlir/test/Pass/ir-printing.mlir

-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-before=cse  -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-before-all -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE_ALL %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-after=cse -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-after-all -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_ALL %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-before=cse -print-ir-module-scope -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE_MODULE %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,cse)' -print-ir-after-all -print-ir-after-change -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_ALL_CHANGE %s
-// RUN: not mlir-opt %s -mlir-disable-threading=true -pass-pipeline='func(cse,test-pass-failure)' -print-ir-after-failure -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_FAILURE %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-before=cse  -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-before-all -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE_ALL %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-after=cse -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-after-all -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_ALL %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-before=cse -print-ir-module-scope -o /dev/null 2>&1 | FileCheck -check-prefix=BEFORE_MODULE %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,cse)' -print-ir-after-all -print-ir-after-change -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_ALL_CHANGE %s
+// RUN: not mlir-opt %s -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,test-pass-failure)' -print-ir-after-failure -o /dev/null 2>&1 | FileCheck -check-prefix=AFTER_FAILURE %s
 
 func @foo() {
   %0 = constant 0 : i32
   return
 }
 
 func @bar() {
   return

 // AFTER_ALL-NEXT: func @foo()
 // AFTER_ALL: // -----// IR Dump After{{.*}}Canonicalizer //----- //
 // AFTER_ALL-NEXT: func @foo()
 // AFTER_ALL: // -----// IR Dump After{{.*}}CSE //----- //
 // AFTER_ALL-NEXT: func @bar()
 // AFTER_ALL: // -----// IR Dump After{{.*}}Canonicalizer //----- //
 // AFTER_ALL-NEXT: func @bar()
 
-// BEFORE_MODULE: // -----// IR Dump Before{{.*}}CSE ('func' operation: @foo) //----- //
+// BEFORE_MODULE: // -----// IR Dump Before{{.*}}CSE ('builtin.func' operation: @foo) //----- //
 // BEFORE_MODULE: func @foo()
 // BEFORE_MODULE: func @bar()
-// BEFORE_MODULE: // -----// IR Dump Before{{.*}}CSE ('func' operation: @bar) //----- //
+// BEFORE_MODULE: // -----// IR Dump Before{{.*}}CSE ('builtin.func' operation: @bar) //----- //
 // BEFORE_MODULE: func @foo()
 // BEFORE_MODULE: func @bar()
 
 // AFTER_ALL_CHANGE: // -----// IR Dump After{{.*}}CSE //----- //
 // AFTER_ALL_CHANGE-NEXT: func @foo()
 // AFTER_ALL_CHANGE-NOT: // -----// IR Dump After{{.*}}CSE //----- //
 // We expect that only 'foo' changed during CSE, and the second run of CSE did
 // nothing.
 
 // AFTER_FAILURE-NOT: // -----// IR Dump After{{.*}}CSE
 // AFTER_FAILURE: // -----// IR Dump After{{.*}}TestFailurePass Failed //----- //

mlir/test/Pass/pass-timing.mlir

-// RUN: mlir-opt %s -mlir-disable-threading=true -verify-each=true -pass-pipeline='func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=list 2>&1 | FileCheck -check-prefix=LIST %s
-// RUN: mlir-opt %s -mlir-disable-threading=true -verify-each=true -pass-pipeline='func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck -check-prefix=PIPELINE %s
-// RUN: mlir-opt %s -mlir-disable-threading=false -verify-each=true -pass-pipeline='func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=list 2>&1 | FileCheck -check-prefix=MT_LIST %s
-// RUN: mlir-opt %s -mlir-disable-threading=false -verify-each=true -pass-pipeline='func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck -check-prefix=MT_PIPELINE %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -verify-each=true -pass-pipeline='builtin.func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=list 2>&1 | FileCheck -check-prefix=LIST %s
+// RUN: mlir-opt %s -mlir-disable-threading=true -verify-each=true -pass-pipeline='builtin.func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck -check-prefix=PIPELINE %s
+// RUN: mlir-opt %s -mlir-disable-threading=false -verify-each=true -pass-pipeline='builtin.func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=list 2>&1 | FileCheck -check-prefix=MT_LIST %s
+// RUN: mlir-opt %s -mlir-disable-threading=false -verify-each=true -pass-pipeline='builtin.func(cse,canonicalize,cse)' -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck -check-prefix=MT_PIPELINE %s
 // RUN: mlir-opt %s -mlir-disable-threading=true -verify-each=false -test-pm-nested-pipeline -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck -check-prefix=NESTED_PIPELINE %s
 
 // LIST: Execution time report
 // LIST: Total Execution Time:
 // LIST: Name
 // LIST-DAG: Canonicalizer
 // LIST-DAG: CSE
 // LIST-DAG: DominanceInfo
 // LIST: Total
 
 // PIPELINE: Execution time report
 // PIPELINE: Total Execution Time:
 // PIPELINE: Name
 // PIPELINE-NEXT: Parser
-// PIPELINE-NEXT: 'func' Pipeline
+// PIPELINE-NEXT: 'builtin.func' Pipeline
 // PIPELINE-NEXT:   CSE
 // PIPELINE-NEXT:     (A) DominanceInfo
 // PIPELINE-NEXT:   Canonicalizer
 // PIPELINE-NEXT:   CSE
 // PIPELINE-NEXT:     (A) DominanceInfo
 // PIPELINE-NEXT: Output
 // PIPELINE-NEXT: Rest
 // PIPELINE-NEXT: Total
 
 // MT_LIST: Execution time report
 // MT_LIST: Total Execution Time:
 // MT_LIST: Name
 // MT_LIST-DAG: Canonicalizer
 // MT_LIST-DAG: CSE
 // MT_LIST-DAG: DominanceInfo
 // MT_LIST: Total
 
 // MT_PIPELINE: Execution time report
 // MT_PIPELINE: Total Execution Time:
 // MT_PIPELINE: Name
 // MT_PIPELINE-NEXT: Parser
-// MT_PIPELINE-NEXT: 'func' Pipeline
+// MT_PIPELINE-NEXT: 'builtin.func' Pipeline
 // MT_PIPELINE-NEXT:   CSE
 // MT_PIPELINE-NEXT:     (A) DominanceInfo
 // MT_PIPELINE-NEXT:   Canonicalizer
 // MT_PIPELINE-NEXT:   CSE
 // MT_PIPELINE-NEXT:     (A) DominanceInfo
 // MT_PIPELINE-NEXT: Output
 // MT_PIPELINE-NEXT: Rest
 // MT_PIPELINE-NEXT: Total
 
 // NESTED_PIPELINE: Execution time report
 // NESTED_PIPELINE: Total Execution Time:
 // NESTED_PIPELINE: Name
 // NESTED_PIPELINE-NEXT: Parser
-// NESTED_PIPELINE-NEXT: Pipeline Collection : ['func', 'module']
-// NESTED_PIPELINE-NEXT:   'func' Pipeline
+// NESTED_PIPELINE-NEXT: Pipeline Collection : ['builtin.func', 'builtin.module']
+// NESTED_PIPELINE-NEXT:   'builtin.func' Pipeline
 // NESTED_PIPELINE-NEXT:     TestFunctionPass
-// NESTED_PIPELINE-NEXT:   'module' Pipeline
+// NESTED_PIPELINE-NEXT:   'builtin.module' Pipeline
 // NESTED_PIPELINE-NEXT:     TestModulePass
-// NESTED_PIPELINE-NEXT:     'func' Pipeline
+// NESTED_PIPELINE-NEXT:     'builtin.func' Pipeline
 // NESTED_PIPELINE-NEXT:       TestFunctionPass
 // NESTED_PIPELINE-NEXT: Output
 // NESTED_PIPELINE-NEXT: Rest
 // NESTED_PIPELINE-NEXT: Total
 
 func @foo() {
   return
 }

mlir/test/Pass/pipeline-options-parsing.mlir

-// RUN: not mlir-opt %s -pass-pipeline='module(test-module-pass{)' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_1 %s
-// RUN: not mlir-opt %s -pass-pipeline='module(test-module-pass{test-option=3})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_2 %s
-// RUN: not mlir-opt %s -pass-pipeline='module(func(test-options-pass{list=3}), test-module-pass{invalid-option=3})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_3 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(test-module-pass{)' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_1 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(test-module-pass{test-option=3})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_2 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(builtin.func(test-options-pass{list=3}), test-module-pass{invalid-option=3})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_3 %s
 // RUN: not mlir-opt %s -pass-pipeline='test-options-pass{list=3 list=notaninteger}' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_4 %s
-// RUN: not mlir-opt %s -pass-pipeline='func(test-options-pass{list=1,2,3,4 list=5 string=value1 string=value2})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_5 %s
-// RUN: mlir-opt %s -verify-each=false -pass-pipeline='func(test-options-pass{string-list=a list=1,2,3,4 string-list=b,c list=5 string-list=d string=some_value})' -test-dump-pipeline 2>&1 | FileCheck --check-prefix=CHECK_1 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.func(test-options-pass{list=1,2,3,4 list=5 string=value1 string=value2})' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_5 %s
+// RUN: mlir-opt %s -verify-each=false -pass-pipeline='builtin.func(test-options-pass{string-list=a list=1,2,3,4 string-list=b,c list=5 string-list=d string=some_value})' -test-dump-pipeline 2>&1 | FileCheck --check-prefix=CHECK_1 %s
 // RUN: mlir-opt %s -verify-each=false -test-options-pass-pipeline='list=1 string-list=a,b' -test-dump-pipeline 2>&1 | FileCheck --check-prefix=CHECK_2 %s
-// RUN: mlir-opt %s -verify-each=false -pass-pipeline='module(func(test-options-pass{list=3}), func(test-options-pass{list=1,2,3,4}))' -test-dump-pipeline 2>&1 | FileCheck --check-prefix=CHECK_3 %s
+// RUN: mlir-opt %s -verify-each=false -pass-pipeline='builtin.module(builtin.func(test-options-pass{list=3}), builtin.func(test-options-pass{list=1,2,3,4}))' -test-dump-pipeline 2>&1 | FileCheck --check-prefix=CHECK_3 %s
 
 // CHECK_ERROR_1: missing closing '}' while processing pass options
 // CHECK_ERROR_2: no such option test-option
 // CHECK_ERROR_3: no such option invalid-option
 // CHECK_ERROR_4: 'notaninteger' value invalid for integer argument
 // CHECK_ERROR_5: string option: may only occur zero or one times
 
 // CHECK_1: test-options-pass{list=1,2,3,4,5 string=some_value string-list=a,b,c,d}
 // CHECK_2: test-options-pass{list=1 string= string-list=a,b}
-// CHECK_3: module(func(test-options-pass{list=3 string= }), func(test-options-pass{list=1,2,3,4 string= }))
+// CHECK_3: builtin.module(builtin.func(test-options-pass{list=3 string= }), builtin.func(test-options-pass{list=1,2,3,4 string= }))

mlir/test/Pass/pipeline-parsing.mlir

-// RUN: mlir-opt %s -mlir-disable-threading -pass-pipeline='module(test-module-pass,func(test-function-pass)),func(test-function-pass)' -pass-pipeline="func(cse,canonicalize)" -verify-each=false -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck %s
+// RUN: mlir-opt %s -mlir-disable-threading -pass-pipeline='builtin.module(test-module-pass,builtin.func(test-function-pass)),builtin.func(test-function-pass)' -pass-pipeline="builtin.func(cse,canonicalize)" -verify-each=false -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck %s
 // RUN: mlir-opt %s -mlir-disable-threading -test-textual-pm-nested-pipeline -verify-each=false -mlir-timing -mlir-timing-display=tree 2>&1 | FileCheck %s --check-prefix=TEXTUAL_CHECK
-// RUN: not mlir-opt %s -pass-pipeline='module(test-module-pass' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_1 %s
-// RUN: not mlir-opt %s -pass-pipeline='module(test-module-pass))' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_2 %s
-// RUN: not mlir-opt %s -pass-pipeline='module()(' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_3 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(test-module-pass' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_1 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module(test-module-pass))' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_2 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.module()(' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_3 %s
 // RUN: not mlir-opt %s -pass-pipeline=',' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_4 %s
-// RUN: not mlir-opt %s -pass-pipeline='func(test-module-pass)' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_5 %s
+// RUN: not mlir-opt %s -pass-pipeline='builtin.func(test-module-pass)' 2>&1 | FileCheck --check-prefix=CHECK_ERROR_5 %s
 
 // CHECK_ERROR_1: encountered unbalanced parentheses while parsing pipeline
 // CHECK_ERROR_2: encountered extra closing ')' creating unbalanced parentheses while parsing pipeline
 // CHECK_ERROR_3: expected ',' after parsing pipeline
 // CHECK_ERROR_4: does not refer to a registered pass or pass pipeline
-// CHECK_ERROR_5:  Can't add pass '{{.*}}TestModulePass' restricted to 'module' on a PassManager intended to run on 'func', did you intend to nest?
+// CHECK_ERROR_5:  Can't add pass '{{.*}}TestModulePass' restricted to 'builtin.module' on a PassManager intended to run on 'builtin.func', did you intend to nest?
 func @foo() {
   return
 }
 
 module {
   func @foo() {
     return
   }
 }
 
-// CHECK: Pipeline Collection : ['func', 'module']
-// CHECK-NEXT:   'func' Pipeline
+// CHECK: Pipeline Collection : ['builtin.func', 'builtin.module']
+// CHECK-NEXT:   'builtin.func' Pipeline
 // CHECK-NEXT:     TestFunctionPass
 // CHECK-NEXT:     CSE
 // CHECK-NEXT:       DominanceInfo
 // CHECK-NEXT:     Canonicalizer
-// CHECK-NEXT:   'module' Pipeline
+// CHECK-NEXT:   'builtin.module' Pipeline
 // CHECK-NEXT:     TestModulePass
-// CHECK-NEXT:     'func' Pipeline
+// CHECK-NEXT:     'builtin.func' Pipeline
 // CHECK-NEXT:       TestFunctionPass
 
-// TEXTUAL_CHECK: Pipeline Collection : ['func', 'module']
-// TEXTUAL_CHECK-NEXT:   'func' Pipeline
+// TEXTUAL_CHECK: Pipeline Collection : ['builtin.func', 'builtin.module']
+// TEXTUAL_CHECK-NEXT:   'builtin.func' Pipeline
 // TEXTUAL_CHECK-NEXT:     TestFunctionPass
-// TEXTUAL_CHECK-NEXT:   'module' Pipeline
+// TEXTUAL_CHECK-NEXT:   'builtin.module' Pipeline
 // TEXTUAL_CHECK-NEXT:     TestModulePass
-// TEXTUAL_CHECK-NEXT:     'func' Pipeline
+// TEXTUAL_CHECK-NEXT:     'builtin.func' Pipeline
 // TEXTUAL_CHECK-NEXT:       TestFunctionPass

mlir/test/Pass/pipeline-stats.mlir

 // REQUIRES: asserts
-// RUN: mlir-opt %s -verify-each=true -pass-pipeline='func(test-stats-pass,test-stats-pass)' -pass-statistics -pass-statistics-display=list 2>&1 | FileCheck -check-prefix=LIST %s
-// RUN: mlir-opt %s -verify-each=true -pass-pipeline='func(test-stats-pass,test-stats-pass)' -pass-statistics -pass-statistics-display=pipeline 2>&1 | FileCheck -check-prefix=PIPELINE %s
+// RUN: mlir-opt %s -verify-each=true -pass-pipeline='builtin.func(test-stats-pass,test-stats-pass)' -pass-statistics -pass-statistics-display=list 2>&1 | FileCheck -check-prefix=LIST %s
+// RUN: mlir-opt %s -verify-each=true -pass-pipeline='builtin.func(test-stats-pass,test-stats-pass)' -pass-statistics -pass-statistics-display=pipeline 2>&1 | FileCheck -check-prefix=PIPELINE %s
 
 // LIST: Pass statistics report
 // LIST: TestStatisticPass
 // LIST-NEXT:  (S) {{0|8}} num-ops - Number of operations counted
 // LIST-NOT: Verifier
 
 // PIPELINE: Pass statistics report
-// PIPELINE: 'func' Pipeline
+// PIPELINE: 'builtin.func' Pipeline
 // PIPELINE-NEXT:   TestStatisticPass
 // PIPELINE-NEXT:     (S) {{0|4}} num-ops - Number of operations counted
 // PIPELINE-NEXT:   TestStatisticPass
 // PIPELINE-NEXT:     (S) {{0|4}} num-ops - Number of operations counted
 
 func @foo() {
   return
 }
 
 func @bar() {
   return
 }

mlir/test/Pass/run-reproducer.mlir

-// configuration: -mlir-disable-threading=true -pass-pipeline='func(cse,canonicalize)' -print-ir-before=cse
+// configuration: -mlir-disable-threading=true -pass-pipeline='builtin.func(cse,canonicalize)' -print-ir-before=cse
 
 // Test of the reproducer run option. The first line has to be the
 // configuration (matching what is produced by reproducer).
 
 // RUN: mlir-opt %s -run-reproducer 2>&1 | FileCheck -check-prefix=BEFORE %s
 
 func @foo() {
   %0 = constant 0 : i32

mlir/test/Transforms/canonicalize-block-merge.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(canonicalize)' -split-input-file | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(canonicalize)' -split-input-file | FileCheck %s
 
 // Check the simple case of single operation blocks with a return.
 
 // CHECK-LABEL: func @return_blocks(
 func @return_blocks() {
   // CHECK: "foo.cond_br"()[^bb1, ^bb1]
   // CHECK: ^bb1:
   // CHECK-NEXT: return

mlir/test/Transforms/canonicalize-dce.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -split-input-file -pass-pipeline='func(canonicalize)' | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -split-input-file -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s
 
 // Test case: Simple case of deleting a dead pure op.
 
 // CHECK:      func @f(%arg0: f32) {
 // CHECK-NEXT:   return
 
 func @f(%arg0: f32) {
   %0 = "std.addf"(%arg0, %arg0) : (f32, f32) -> f32

mlir/test/Transforms/canonicalize-td.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(canonicalize{top-down=true})' | FileCheck %s --check-prefix=TD
-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(canonicalize)' | FileCheck %s --check-prefix=BU
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(canonicalize{top-down=true})' | FileCheck %s --check-prefix=TD
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s --check-prefix=BU
 
 
 // BU-LABEL: func @default_insertion_position
 // TD-LABEL: func @default_insertion_position
 func @default_insertion_position(%cond: i1) {
   // Constant should be folded into the entry block.
 
   // BU: constant 2

mlir/test/Transforms/canonicalize.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(canonicalize)' -split-input-file | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(canonicalize)' -split-input-file | FileCheck %s
 
 // CHECK-LABEL: func @test_subi_zero
 func @test_subi_zero(%arg0: i32) -> i32 {
   // CHECK-NEXT: %c0_i32 = constant 0 : i32
   // CHECK-NEXT: return %c0
   %y = subi %arg0, %arg0 : i32
   return %y: i32
 }

mlir/test/Transforms/cse.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(cse)' | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(cse)' | FileCheck %s
 
 // CHECK-DAG: #[[$MAP:.*]] = affine_map<(d0) -> (d0 mod 2)>
 #map0 = affine_map<(d0) -> (d0 mod 2)>
 
 // CHECK-LABEL: @simple_constant
 func @simple_constant() -> (i32, i32) {
   // CHECK-NEXT: %c1_i32 = constant 1 : i32
   %0 = constant 1 : i32

mlir/test/Transforms/parallel-loop-collapsing.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(parallel-loop-collapsing{collapsed-indices-0=0,3 collapsed-indices-1=1,4 collapsed-indices-2=2}, canonicalize)' | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(parallel-loop-collapsing{collapsed-indices-0=0,3 collapsed-indices-1=1,4 collapsed-indices-2=2}, canonicalize)' | FileCheck %s
 
 // CHECK-LABEL: func @parallel_many_dims() {
 func @parallel_many_dims() {
   %c0 = constant 0 : index
   %c1 = constant 1 : index
   %c2 = constant 2 : index
   %c3 = constant 3 : index
   %c4 = constant 4 : index

mlir/test/Transforms/sccp-callgraph.mlir

 // RUN: mlir-opt -allow-unregistered-dialect %s -sccp -split-input-file | FileCheck %s
-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="module(sccp)" -split-input-file | FileCheck %s --check-prefix=NESTED
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="builtin.module(sccp)" -split-input-file | FileCheck %s --check-prefix=NESTED
 
 /// Check that a constant is properly propagated through the arguments and
 /// results of a private function.
 
 // CHECK-LABEL: func private @private(
 func private @private(%arg0 : i32) -> i32 {
   // CHECK: %[[CST:.*]] = constant 1 : i32
   // CHECK: return %[[CST]] : i32

mlir/test/Transforms/sccp-structured.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="func(sccp)" -split-input-file | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="builtin.func(sccp)" -split-input-file | FileCheck %s
 
 /// Check that a constant is properly propagated when only one edge is taken.
 
 // CHECK-LABEL: func @simple(
 func @simple(%arg0 : i32) -> i32 {
   // CHECK: %[[CST:.*]] = constant 1 : i32
   // CHECK-NOT: scf.if
   // CHECK: return %[[CST]] : i32

mlir/test/Transforms/sccp.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="func(sccp)" -split-input-file | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="builtin.func(sccp)" -split-input-file | FileCheck %s
 
 /// Check simple forward constant propagation without any control flow.
 
 // CHECK-LABEL: func @no_control_flow
 func @no_control_flow(%arg0: i32) -> i32 {
   // CHECK: %[[CST:.*]] = constant 1 : i32
   // CHECK: return %[[CST]] : i32
 

mlir/test/Transforms/single-parallel-loop-collapsing.mlir

-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='func(parallel-loop-collapsing{collapsed-indices-0=0,1}, canonicalize)' | FileCheck %s
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline='builtin.func(parallel-loop-collapsing{collapsed-indices-0=0,1}, canonicalize)' | FileCheck %s
 
 func @collapse_to_single() {
   %c0 = constant 3 : index
   %c1 = constant 7 : index
   %c2 = constant 11 : index
   %c3 = constant 29 : index
   %c4 = constant 3 : index
   %c5 = constant 4 : index

mlir/test/Transforms/test-canonicalize-filter.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize)' | FileCheck %s --check-prefix=NO_FILTER
-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize{enable-patterns=TestRemoveOpWithInnerOps})' | FileCheck %s --check-prefix=FILTER_ENABLE
-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize{disable-patterns=TestRemoveOpWithInnerOps})' | FileCheck %s --check-prefix=FILTER_DISABLE
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s --check-prefix=NO_FILTER
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize{enable-patterns=TestRemoveOpWithInnerOps})' | FileCheck %s --check-prefix=FILTER_ENABLE
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize{disable-patterns=TestRemoveOpWithInnerOps})' | FileCheck %s --check-prefix=FILTER_DISABLE
 
 // NO_FILTER-LABEL: func @remove_op_with_inner_ops_pattern
 // NO_FILTER-NEXT: return
 // FILTER_ENABLE-LABEL: func @remove_op_with_inner_ops_pattern
 // FILTER_ENABLE-NEXT: return
 // FILTER_DISABLE-LABEL: func @remove_op_with_inner_ops_pattern
 // FILTER_DISABLE-NEXT: "test.op_with_region_pattern"()
 func @remove_op_with_inner_ops_pattern() {
   "test.op_with_region_pattern"() ({
     "test.op_with_region_terminator"() : () -> ()
   }) : () -> ()
   return
 }

mlir/test/Transforms/test-canonicalize.mlir

-// RUN: mlir-opt %s -pass-pipeline='func(canonicalize)' | FileCheck %s
+// RUN: mlir-opt %s -pass-pipeline='builtin.func(canonicalize)' | FileCheck %s
 
 // CHECK-LABEL: func @remove_op_with_inner_ops_pattern
 func @remove_op_with_inner_ops_pattern() {
   // CHECK-NEXT: return
   "test.op_with_region_pattern"() ({
     "test.op_with_region_terminator"() : () -> ()
   }) : () -> ()
   return

mlir/test/Transforms/test-legalizer-analysis.mlir

 // RUN: mlir-opt -allow-unregistered-dialect -test-legalize-patterns -verify-diagnostics -test-legalize-mode=analysis %s | FileCheck %s
-// expected-remark@-2 {{op 'module' is legalizable}}
+// expected-remark@-2 {{op 'builtin.module' is legalizable}}
 
-// expected-remark@+1 {{op 'func' is legalizable}}
+// expected-remark@+1 {{op 'builtin.func' is legalizable}}
 func @test(%arg0: f32) {
   // expected-remark@+1 {{op 'test.illegal_op_a' is legalizable}}
   %result = "test.illegal_op_a"() : () -> (i32)
   "foo.region"() ({
       // expected-remark@+1 {{op 'test.invalid' is legalizable}}
       "test.invalid"() : () -> ()
   }) : () -> ()
   return
 }
 
 // Check that none of the legalizable operations were modified.
 // CHECK-LABEL: func @test
 // CHECK-NEXT: "test.illegal_op_a"
 // CHECK: "test.invalid"

mlir/test/Transforms/test-symbol-dce.mlir

 // RUN: mlir-opt -allow-unregistered-dialect %s -symbol-dce -split-input-file -verify-diagnostics | FileCheck %s
-// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="module(symbol-dce)" -split-input-file | FileCheck %s --check-prefix=NESTED
+// RUN: mlir-opt -allow-unregistered-dialect %s -pass-pipeline="builtin.module(symbol-dce)" -split-input-file | FileCheck %s --check-prefix=NESTED
 
 // Check that trivially dead and trivially live non-nested cases are handled.
 
 // CHECK-LABEL: module attributes {test.simple}
 module attributes {test.simple} {
   // CHECK-NOT: func private @dead_private_function
   func private @dead_private_function()
 

mlir/test/mlir-lsp-server/diagnostics.test

 // RUN: mlir-lsp-server -lit-test < %s | FileCheck -strict-whitespace %s
 {"jsonrpc":"2.0","id":0,"method":"initialize","params":{"processId":123,"rootPath":"mlir","capabilities":{},"trace":"off"}}
 // -----
 {"jsonrpc":"2.0","method":"textDocument/didOpen","params":{"textDocument":{
   "uri":"test:///foo.mlir",
   "languageId":"mlir",
   "version":1,
   "text":"func ()"
 }}}
 // CHECK: "method": "textDocument/publishDiagnostics",
 // CHECK-NEXT: "params": {
 // CHECK-NEXT:     "diagnostics": [
 // CHECK-NEXT:       {
 // CHECK-NEXT:         "category": "Parse Error",
-// CHECK-NEXT:         "message": "custom op 'func' expected valid '@'-identifier for symbol name",
+// CHECK-NEXT:         "message": "custom op 'builtin.func' expected valid '@'-identifier for symbol name",
 // CHECK-NEXT:         "range": {
 // CHECK-NEXT:           "end": {
 // CHECK-NEXT:             "character": 7,
 // CHECK-NEXT:             "line": 0
 // CHECK-NEXT:           },
 // CHECK-NEXT:           "start": {
 // CHECK-NEXT:             "character": 6,
 // CHECK-NEXT:             "line": 0

mlir/test/mlir-lsp-server/document-symbols.test

 // CHECK-NEXT:            "start": {
 // CHECK-NEXT:              "character": {{.*}},
 // CHECK-NEXT:              "line": {{.*}}
 // CHECK-NEXT:            }
 // CHECK-NEXT:          }
 // CHECK-NEXT:        }
 // CHECK-NEXT:      ],
 // CHECK-NEXT:      "kind": 3,
-// CHECK-NEXT:      "name": "<module>",
+// CHECK-NEXT:      "name": "<builtin.module>",
 // CHECK-NEXT:      "range": {
 // CHECK-NEXT:        "end": {
 // CHECK-NEXT:          "character": {{.*}},
 // CHECK-NEXT:          "line": {{.*}}
 // CHECK-NEXT:        },
 // CHECK-NEXT:        "start": {
 // CHECK-NEXT:          "character": {{.*}},
 // CHECK-NEXT:          "line": {{.*}}

mlir/test/mlir-lsp-server/hover.test

// Hover on a Block.
   "textDocument":{"uri":"test:///foo.mlir"},
   "position":{"line":3,"character":2}
 }}
 //      CHECK:  "id": 3,
 // CHECK-NEXT:  "jsonrpc": "2.0",
 // CHECK-NEXT:  "result": {
 // CHECK-NEXT:    "contents": {
 // CHECK-NEXT:      "kind": "markdown",
-// CHECK-NEXT:      "value": "Operation: \"func\"\n\nBlock #1\n\nPredecessors: <Block #0>\n\n"
+// CHECK-NEXT:      "value": "Operation: \"builtin.func\"\n\nBlock #1\n\nPredecessors: <Block #0>\n\n"
 // CHECK-NEXT:    },
 // CHECK-NEXT:    "range": {
 // CHECK-NEXT:      "end": {
 // CHECK-NEXT:        "character": 4,
 // CHECK-NEXT:        "line": 3
 // CHECK-NEXT:      },
 // CHECK-NEXT:      "start": {
 // CHECK-NEXT:        "character": 0,
 // CHECK-NEXT:        "line": 3
 // CHECK-NEXT:      }
 // CHECK-NEXT:    }
 // CHECK-NEXT:  }
 // -----
 // Hover on a Block argument.
 {"jsonrpc":"2.0","id":4,"method":"textDocument/hover","params":{
   "textDocument":{"uri":"test:///foo.mlir"},
   "position":{"line":0,"character":12}
 }}
 //      CHECK:  "id": 4,
 // CHECK-NEXT:  "jsonrpc": "2.0",
 // CHECK-NEXT:  "result": {
 // CHECK-NEXT:    "contents": {
 // CHECK-NEXT:      "kind": "markdown",
-// CHECK-NEXT:      "value": "Operation: \"func\"\n\nBlock: <Block #0>\n\nArgument #0\n\nType: `i1`\n\n"
+// CHECK-NEXT:      "value": "Operation: \"builtin.func\"\n\nBlock: <Block #0>\n\nArgument #0\n\nType: `i1`\n\n"
 // CHECK-NEXT:    },
 // CHECK-NEXT:    "range": {
 // CHECK-NEXT:      "end": {
 // CHECK-NEXT:        "character": 14,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      },
 // CHECK-NEXT:      "start": {
 // CHECK-NEXT:        "character": 10,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      }
 // CHECK-NEXT:    }
 // CHECK-NEXT:  }
 // -----
 // Hover on a region operation.
 {"jsonrpc":"2.0","id":5,"method":"textDocument/hover","params":{
   "textDocument":{"uri":"test:///foo.mlir"},
   "position":{"line":0,"character":1}
 }}
 //      CHECK:  "id": 5,
 // CHECK-NEXT:  "jsonrpc": "2.0",
 // CHECK-NEXT:  "result": {
 // CHECK-NEXT:    "contents": {
 // CHECK-NEXT:      "kind": "markdown",
-// CHECK-NEXT:      "value": "\"func\" : public @foo\n\nGeneric Form:\n\n```mlir\n\"func\"() ( {\n}) {sym_name = \"foo\", type = (i1) -> ()} : () -> ()\n```\n"
+// CHECK-NEXT:      "value": "\"builtin.func\" : public @foo\n\nGeneric Form:\n\n```mlir\n\"builtin.func\"() ( {\n}) {sym_name = \"foo\", type = (i1) -> ()} : () -> ()\n```\n"
 // CHECK-NEXT:    },
 // CHECK-NEXT:    "range": {
 // CHECK-NEXT:      "end": {
 // CHECK-NEXT:        "character": 4,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      },
 // CHECK-NEXT:      "start": {
 // CHECK-NEXT:        "character": 0,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      }
 // CHECK-NEXT:    }
 // CHECK-NEXT:  }
 // -----
 // Hover on a symbol reference.
 {"jsonrpc":"2.0","id":6,"method":"textDocument/hover","params":{
   "textDocument":{"uri":"test:///foo.mlir"},
   "position":{"line":0,"character":8}
 }}
 //      CHECK:  "id": 6,
 // CHECK-NEXT:  "jsonrpc": "2.0",
 // CHECK-NEXT:  "result": {
 // CHECK-NEXT:    "contents": {
 // CHECK-NEXT:      "kind": "markdown",
-// CHECK-NEXT:      "value": "\"func\" : public @foo\n\nGeneric Form:\n\n```mlir\n\"func\"() ( {\n}) {sym_name = \"foo\", type = (i1) -> ()} : () -> ()\n```\n"
+// CHECK-NEXT:      "value": "\"builtin.func\" : public @foo\n\nGeneric Form:\n\n```mlir\n\"builtin.func\"() ( {\n}) {sym_name = \"foo\", type = (i1) -> ()} : () -> ()\n```\n"
 // CHECK-NEXT:    },
 // CHECK-NEXT:    "range": {
 // CHECK-NEXT:      "end": {
 // CHECK-NEXT:        "character": 9,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      },
 // CHECK-NEXT:      "start": {
 // CHECK-NEXT:        "character": 5,
 // CHECK-NEXT:        "line": 0
 // CHECK-NEXT:      }
 // CHECK-NEXT:    }
 // CHECK-NEXT:  }
 // -----
 {"jsonrpc":"2.0","id":7,"method":"shutdown"}
 // -----
 {"jsonrpc":"2.0","method":"exit"}

mlir/test/mlir-opt/commandline.mlir

 // RUN: mlir-opt --show-dialects | FileCheck %s
 // CHECK: Available Dialects:
 // CHECK-NEXT: acc
 // CHECK-NEXT: affine
 // CHECK-NEXT: amx
 // CHECK-NEXT: arm_neon
 // CHECK-NEXT: arm_sve
 // CHECK-NEXT: async
+// CHECK-NEXT: builtin
 // CHECK-NEXT: complex
 // CHECK-NEXT: dlti
 // CHECK-NEXT: emitc
 // CHECK-NEXT: gpu
 // CHECK-NEXT: linalg
 // CHECK-NEXT: llvm
 // CHECK-NEXT: math
 // CHECK-NEXT: memref

mlir/test/python/integration/dialects/linalg/opsrun.py

def transform(module, boilerplate):
   import mlir.dialects.linalg.passes
   import mlir.transforms
 
   # TODO: Allow cloning functions from one module to another.
   # Atm we have to resort to string concatenation.
   mod = Module.parse(
       str(module.operation.regions[0].blocks[0].operations[0].operation) +
       boilerplate)
-  pm = PassManager.parse("func(convert-linalg-to-loops, lower-affine, " +
-                         "convert-scf-to-std), convert-vector-to-llvm," +
-                         "convert-memref-to-llvm,convert-std-to-llvm")
+  pm = PassManager.parse(
+      "builtin.func(convert-linalg-to-loops, lower-affine, " +
   pm.run(mod)
   return mod
 
 
 def test_matmul_builtin():
   with Context() as ctx, Location.unknown():
     module = Module.create()
     f32 = F32Type.get()

mlir/test/python/ir/dialects.py

 
 
 # CHECK-LABEL: TEST: testUserDialectClass
 @run
 def testUserDialectClass():
   ctx = Context()
   # Access using attribute.
   d = ctx.dialects.std
-  # Note that the standard dialect namespace prints as ''. Others will print
-  # as "<Dialect %namespace (..."
-  # CHECK: <Dialect (class mlir.dialects._std_ops_gen._Dialect)>
+  # CHECK: <Dialect std (class mlir.dialects._std_ops_gen._Dialect)>
   print(d)
   try:
     _ = ctx.dialects.not_existing
   except AttributeError:
     pass
   else:
     assert False, "Expected exception"
 
   # Access using index.
   d = ctx.dialects["std"]
-  # CHECK: <Dialect (class mlir.dialects._std_ops_gen._Dialect)>
+  # CHECK: <Dialect std (class mlir.dialects._std_ops_gen._Dialect)>
   print(d)
   try:
     _ = ctx.dialects["not_existing"]
   except IndexError:
     pass
   else:
     assert False, "Expected exception"
 
   # Using the 'd' alias.
   d = ctx.d["std"]
-  # CHECK: <Dialect (class mlir.dialects._std_ops_gen._Dialect)>
+  # CHECK: <Dialect std (class mlir.dialects._std_ops_gen._Dialect)>
   print(d)
 
 
 # CHECK-LABEL: TEST: testCustomOpView
 # This test uses the standard dialect AddFOp as an example of a user op.
 # TODO: Op creation and access is still quite verbose: simplify this test as
 # additional capabilities come online.
 @run

mlir/test/python/ir/operation.py

def testSingleResultProperty():
   print(module.body.operations[2])
 
 run(testSingleResultProperty)
 
 # CHECK-LABEL: TEST: testPrintInvalidOperation
 def testPrintInvalidOperation():
   ctx = Context()
   with Location.unknown(ctx):
-    module = Operation.create("module", regions=2)
+    module = Operation.create("builtin.module", regions=2)
     # This module has two region and is invalid verify that we fallback
     # to the generic printer for safety.
     block = module.regions[0].blocks.append()
     # CHECK: // Verification failed, printing generic form
     # CHECK: "module"() ( {
     # CHECK: }) : () -> ()
     print(module)
     # CHECK: .verify = False
     print(f".verify = {module.operation.verify()}")
 run(testPrintInvalidOperation)
 
 
 # CHECK-LABEL: TEST: testCreateWithInvalidAttributes
 def testCreateWithInvalidAttributes():
   ctx = Context()
   with Location.unknown(ctx):
     try:
-      Operation.create("module", attributes={None:StringAttr.get("name")})
+      Operation.create("builtin.module", attributes={None:StringAttr.get("name")})
     except Exception as e:
       # CHECK: Invalid attribute key (not a string) when attempting to create the operation "module"
       print(e)
     try:
-      Operation.create("module", attributes={42:StringAttr.get("name")})
+      Operation.create("builtin.module", attributes={42:StringAttr.get("name")})
     except Exception as e:
       # CHECK: Invalid attribute key (not a string) when attempting to create the operation "module"
       print(e)
     try:
-      Operation.create("module", attributes={"some_key":ctx})
+      Operation.create("builtin.module", attributes={"some_key":ctx})
     except Exception as e:
       # CHECK: Invalid attribute value for the key "some_key" when attempting to create the operation "module"
       print(e)
     try:
-      Operation.create("module", attributes={"some_key":None})
+      Operation.create("builtin.module", attributes={"some_key":None})
     except Exception as e:
       # CHECK: Found an invalid (`None`?) attribute value for the key "some_key" when attempting to create the operation "module"
       print(e)
 run(testCreateWithInvalidAttributes)
 
 
 # CHECK-LABEL: TEST: testOperationName
 def testOperationName():

mlir/test/python/pass_manager.py

 
 # Verify successful round-trip.
 # CHECK-LABEL: TEST: testParseSuccess
 def testParseSuccess():
   with Context():
     # A first import is expected to fail because the pass isn't registered
     # until we import mlir.transforms
     try:
-      pm = PassManager.parse("module(func(print-op-stats))")
+      pm = PassManager.parse("builtin.module(builtin.func(print-op-stats))")
       # TODO: this error should be propagate to Python but the C API does not help right now.
       # CHECK: error: 'print-op-stats' does not refer to a registered pass or pass pipeline
     except ValueError as e:
-      # CHECK: ValueError exception: invalid pass pipeline 'module(func(print-op-stats))'.
+      # CHECK: ValueError exception: invalid pass pipeline 'builtin.module(builtin.func(print-op-stats))'.
       log("ValueError exception:", e)
     else:
       log("Exception not produced")
 
     # This will register the pass and round-trip should be possible now.
     import mlir.transforms
-    pm = PassManager.parse("module(func(print-op-stats))")
-    # CHECK: Roundtrip: module(func(print-op-stats))
+    pm = PassManager.parse("builtin.module(builtin.func(print-op-stats))")
+    # CHECK: Roundtrip: builtin.module(builtin.func(print-op-stats))
     log("Roundtrip: ", pm)
 run(testParseSuccess)
 
 # Verify failure on unregistered pass.
 # CHECK-LABEL: TEST: testParseFail
 def testParseFail():
   with Context():
     try:
       pm = PassManager.parse("unknown-pass")
     except ValueError as e:
       # CHECK: ValueError exception: invalid pass pipeline 'unknown-pass'.
       log("ValueError exception:", e)
     else:
       log("Exception not produced")
 run(testParseFail)
 
 
 # Verify failure on incorrect level of nesting.
 # CHECK-LABEL: TEST: testInvalidNesting
 def testInvalidNesting():
   with Context():
     try:
-      pm = PassManager.parse("func(view-op-graph)")
+      pm = PassManager.parse("builtin.func(view-op-graph)")
     except ValueError as e:
-      # CHECK: Can't add pass 'ViewOpGraphPass' restricted to 'module' on a PassManager intended to run on 'func', did you intend to nest?
-      # CHECK: ValueError exception: invalid pass pipeline 'func(view-op-graph)'.
+      # CHECK: Can't add pass 'ViewOpGraphPass' restricted to 'builtin.module' on a PassManager intended to run on 'builtin.func', did you intend to nest?
+      # CHECK: ValueError exception: invalid pass pipeline 'builtin.func(view-op-graph)'.
       log("ValueError exception:", e)
     else:
       log("Exception not produced")
 run(testInvalidNesting)
 
 
 # Verify that a pass manager can execute on IR
 # CHECK-LABEL: TEST: testRun
 def testRunPipeline():
   with Context():
     pm = PassManager.parse("print-op-stats")
     module = Module.parse(r"""func @successfulParse() { return }""")
     pm.run(module)
 # CHECK: Operations encountered:
-# CHECK: func              , 1
-# CHECK: module            , 1
+# CHECK: builtin.func      , 1
+# CHECK: builtin.module    , 1
 # CHECK: std.return        , 1
 run(testRunPipeline)

mlir/tools/mlir-tblgen/OpPythonBindingGen.cpp

if (clDialectName.empty())
     llvm::PrintFatalError("dialect name not provided");
 
   AttributeClasses attributeClasses;
   constructAttributeMapping(records, attributeClasses);
 
   os << llvm::formatv(fileHeader, clDialectName.getValue());
   os << llvm::formatv(dialectClassTemplate, clDialectName.getValue());
 
-  if (clDialectName == "builtin")
-    clDialectName = "";
-
   for (const llvm::Record *rec : records.getAllDerivedDefinitions("Op")) {
     Operator op(rec);
     if (op.getDialectName() == clDialectName.getValue())
       emitOpBindings(op, attributeClasses, os);
   }
   return false;
 }
 
 static GenRegistration
     genPythonBindings("gen-python-op-bindings",
                       "Generate Python bindings for MLIR Ops", &emitAllOps);

mlir/unittests/IR/InterfaceAttachmentTest.cpp

TEST(InterfaceAttachment, Operation) {
   // Initially, the operation doesn't have the interface.
   auto moduleOp = ModuleOp::create(UnknownLoc::get(&context));
   ASSERT_FALSE(isa<TestExternalOpInterface>(moduleOp.getOperation()));
 
   // We can attach an external interface and now the operaiton has it.
   ModuleOp::attachInterface<TestExternalOpModel>(context);
   auto iface = dyn_cast<TestExternalOpInterface>(moduleOp.getOperation());
   ASSERT_TRUE(iface != nullptr);
-  EXPECT_EQ(iface.getNameLengthPlusArg(10), 16u);
-  EXPECT_EQ(iface.getNameLengthTimesArg(3), 18u);
-  EXPECT_EQ(iface.getNameLengthPlusArgTwice(18), 42u);
-  EXPECT_EQ(iface.getNameLengthMinusArg(5), 1u);
+  EXPECT_EQ(iface.getNameLengthPlusArg(10), 24u);
+  EXPECT_EQ(iface.getNameLengthTimesArg(3), 42u);
+  EXPECT_EQ(iface.getNameLengthPlusArgTwice(18), 50u);
+  EXPECT_EQ(iface.getNameLengthMinusArg(5), 9u);
 
   // Default implementation can be overridden.
   auto funcOp = FuncOp::create(UnknownLoc::get(&context), "function",
                                FunctionType::get(&context, {}, {}));
   ASSERT_FALSE(isa<TestExternalOpInterface>(funcOp.getOperation()));
   FuncOp::attachInterface<TestExternalOpOverridingModel>(context);
   iface = dyn_cast<TestExternalOpInterface>(funcOp.getOperation());
   ASSERT_TRUE(iface != nullptr);
-  EXPECT_EQ(iface.getNameLengthPlusArg(10), 14u);
+  EXPECT_EQ(iface.getNameLengthPlusArg(10), 22u);
   EXPECT_EQ(iface.getNameLengthTimesArg(0), 42u);
-  EXPECT_EQ(iface.getNameLengthPlusArgTwice(8), 20u);
+  EXPECT_EQ(iface.getNameLengthPlusArgTwice(8), 28u);
   EXPECT_EQ(iface.getNameLengthMinusArg(1000), 21u);
 
   // Another context doesn't have the interfaces registered.
   MLIRContext other;
   auto otherModuleOp = ModuleOp::create(UnknownLoc::get(&other));
   ASSERT_FALSE(isa<TestExternalOpInterface>(otherModuleOp.getOperation()));
 }