### start()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1Timer-members.html

Starts the timer.

```APIDOC
## start()

### Description
Starts the timer.

### Method
inline
```

--------------------------------

### start

Source: https://mlir.llvm.org/doxygen/functions_func_s.html

Starts a timer.

```APIDOC
## start()

### Description
Starts the timer.

### Method
start()

### Parameters
None
```

--------------------------------

### OperandRange::getBeginOperandIndex

Source: https://mlir.llvm.org/doxygen/OperationSupport_8cpp_source.html

Gets the starting operand index for the operand range.

```APIDOC
## getBeginOperandIndex

### Description
Gets the starting operand index for the operand range. Asserts if the range is empty.

### Signature
```cpp
unsigned OperandRange::getBeginOperandIndex() const
```
```

--------------------------------

### Get FileLineCol Location Properties (Python)

Source: https://mlir.llvm.org/doxygen/IRCore_8cpp_source.html

Retrieves properties of a FileLineCol location, including filename, start line, start column, end line, and end column.

```python
c.def_prop_ro(
 "filename",
 [](PyFileLineColLocation &self) {
 return mlirIdentifierStr(
 mlirLocationFileLineColRangeGetFilename(self.get()));
 },
 "Gets the filename from a `FileLineColLoc`.");
```

```python
c.def_prop_ro(
 "start_line",
 [](PyFileLineColLocation &self) {
 return mlirLocationFileLineColRangeGetStartLine(self.get());
 },
 "Gets the start line number from a `FileLineColLoc`.");
```

```python
c.def_prop_ro(
 "start_col",
 [](PyFileLineColLocation &self) {
 return mlirLocationFileLineColRangeGetStartColumn(self.get());
 },
 "Gets the start column number from a `FileLineColLoc`.");
```

```python
c.def_prop_ro(
 "end_line",
 [](PyFileLineColLocation &self) {
 return mlirLocationFileLineColRangeGetEndLine(self.get());
 },
 "Gets the end line number from a `FileLineColLoc`.");
```

```python
c.def_prop_ro(
 "end_col",
 [](PyFileLineColLocation &self) {
 return mlirLocationFileLineColRangeGetEndColumn(self.get());
 },
 "Gets the end column number from a `FileLineColLoc`.");
```

--------------------------------

### initialize

Source: https://mlir.llvm.org/doxygen/classmlir_1_1remark_1_1detail_1_1RemarkEngine.html

Sets up the remark engine with the given output streamer and emitting policy.

```APIDOC
## LogicalResult initialize(std::unique_ptr<MLIRRemarkStreamerBase> streamer, std::unique_ptr<RemarkEmittingPolicyBase> remarkEmittingPolicy, std::string *errMsg)

### Description
Setup the remark engine with the given output path and format.

### Parameters
* **streamer** (std::unique_ptr<MLIRRemarkStreamerBase>) - The streamer to use for outputting remarks.
* **remarkEmittingPolicy** (std::unique_ptr<RemarkEmittingPolicyBase>) - The policy to use for emitting remarks.
* **errMsg** (std::string *) - A pointer to a string to store any error messages.

### Returns
* LogicalResult - Indicates success or failure of the initialization.
```

--------------------------------

### DenseElementsAttr::value_begin

Source: https://mlir.llvm.org/doxygen/mlir_2IR_2BuiltinAttributes_8h_source.html

Gets an iterator of the specified type T to the start of the held element values.

```APIDOC
## auto mlir::DenseElementsAttr::value_begin() const

### Description
Get an iterator of the given type to the start of the held element values.

### Definition
BuiltinAttributes.h:418
```

--------------------------------

### initializeBackendsOnce

Source: https://mlir.llvm.org/doxygen/TargetUtils_8h_source.html

Idempotent helper to register/initialize all backends that LLVM has been configured to support. Only runs the first time it is called.

```APIDOC
## initializeBackendsOnce

### Description
Idempotent helper to register/initialize all backends that LLVM has been configured to support. Only runs the first time it is called.

### Signature
void initializeBackendsOnce()

### Usage
This function is typically called internally to ensure LLVM backends are ready before they are used.
```

--------------------------------

### getStartLine

Source: https://mlir.llvm.org/doxygen/functions_func_g.html

Gets the starting line from a file line column range attribute storage.

```APIDOC
## getStartLine

### Description
Gets the starting line from a file line column range attribute storage.

### Signature
mlir::detail::FileLineColRangeAttrStorage getStartLine()

### Returns
A mlir::detail::FileLineColRangeAttrStorage object.
```

--------------------------------

### Create Vulkan Instance

Source: https://mlir.llvm.org/doxygen/VulkanRuntime_8cpp_source.html

Initializes the Vulkan instance, enabling necessary extensions like KHR_PORTABILITY_ENUMERATION on Apple platforms. Ensure Vulkan SDK is correctly set up.

```cpp
instanceCreateInfo.pNext = nullptr;
instanceCreateInfo.pApplicationInfo = &applicationInfo;
instanceCreateInfo.enabledLayerCount = 0;
instanceCreateInfo.ppEnabledLayerNames = nullptr;

std::vector<const char *> extNames;
#if defined(__APPLE__)
// enumerate MoltenVK for Vulkan 1.0
instanceCreateInfo.flags = VK_INSTANCE_CREATE_ENUMERATE_PORTABILITY_BIT_KHR;
// add KHR portability instance extensions
extNames.push_back(VK_KHR_GET_PHYSICAL_DEVICE_PROPERTIES_2_EXTENSION_NAME);
extNames.push_back(VK_KHR_PORTABILITY_ENUMERATION_EXTENSION_NAME);
#else
instanceCreateInfo.flags = 0;
#endif // __APPLE__
instanceCreateInfo.enabledExtensionCount =
static_cast<uint32_t>(extNames.size());
instanceCreateInfo.ppEnabledExtensionNames = extNames.data();

RETURN_ON_VULKAN_ERROR(
vkCreateInstance(&instanceCreateInfo, nullptr, &instance),
"vkCreateInstance");
return success();
```

--------------------------------

### startEmit()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1sparse__tensor_1_1CodegenEnv.html

Starts the emission process with a specified SparseEmitStrategy.

```APIDOC
## startEmit()

### Description
Starts the emission process with a specified SparseEmitStrategy.

### Method
`void CodegenEnv::startEmit(SparseEmitStrategy _emitStrategy_)`

### Definition
`lib/Dialect/SparseTensor/Transforms/Utils/CodegenEnv.cpp:61`

### References
- `mlir::sparse_tensor::Merger::getDependentLoops()`
- `getLoopNum()`
- `mlir::sparse_tensor::getSparseTensorEncoding()`
- `makeTensorId()`
- `merger()`
- `sortDependentLoops()`
```

--------------------------------

### getStartColumn

Source: https://mlir.llvm.org/doxygen/functions_func_g.html

Gets the starting column from a file line column range attribute storage.

```APIDOC
## getStartColumn

### Description
Gets the starting column from a file line column range attribute storage.

### Signature
mlir::detail::FileLineColRangeAttrStorage getStartColumn()

### Returns
A mlir::detail::FileLineColRangeAttrStorage object.
```

--------------------------------

### getPerfectlyNestedLoops

Source: https://mlir.llvm.org/doxygen/include_2mlir_2Dialect_2SCF_2Utils_2Utils_8h.html

Gets a perfectly nested sequence of loops starting at the root of a loop nest.

```APIDOC
## getPerfectlyNestedLoops

### Description
Get perfectly nested sequence of loops starting at root of loop nest (the first op being another AffineFor, and the second op - a terminator).

### Signature
void | mlir::getPerfectlyNestedLoops (SmallVectorImpl< scf::ForOp > &nestedLoops, scf::ForOp root)
```

--------------------------------

### MlirPdllLspServerMain

Source: https://mlir.llvm.org/doxygen/MlirPdllLspServerMain_8h_source.html

This function is the main entry point for tools like `mlir-pdll-lsp-server` when built as a standalone binary. It takes the command-line arguments and initializes the language server.

```APIDOC
## MlirPdllLspServerMain

### Description
Implementation for tools like `mlir-pdll-lsp-server`.

### Signature
llvm::LogicalResult MlirPdllLspServerMain(int argc, char **argv)

### Parameters
* **argc** (int) - The number of command-line arguments.
* **argv** (char **) - An array of command-line argument strings.
```

--------------------------------

### begin()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1StaticTileOffsetRange.html

Returns an iterator to the beginning of the tile offset range.

```APIDOC
## IteratorTy begin() const
Returns an iterator to the beginning of the tile offset range.
```

--------------------------------

### Get Token Location Range

Source: https://mlir.llvm.org/doxygen/Token_8cpp_source.html

Returns the start and end locations of the token as a range.

```C++
SMRange Token::getLocRange() const { return SMRange(getLoc(), getEndLoc()); }
```

--------------------------------

### init

Source: https://mlir.llvm.org/doxygen/classTuneExtension-members.html

Initializes the TuneExtension.

```APIDOC
## init

### Description
Initializes the TuneExtension.

### Signature
TuneExtension::init()

### Method
Inline
```

--------------------------------

### Get Token Location

Source: https://mlir.llvm.org/doxygen/Token_8cpp_source.html

Retrieves the starting location of the token within the source file.

```C++
SMLoc Token::getLoc() const {
  return SMLoc::getFromPointer(spelling.data());
}
```

--------------------------------

### initializeBackendsOnce

Source: https://mlir.llvm.org/doxygen/TargetUtils_8h.html

Idempotent helper to register and initialize all backends that LLVM has been configured to support.

```APIDOC
## initializeBackendsOnce

### Description
Idempotent helper to register/initialize all backends that LLVM has been configured to support.

### Signature
`void mlir::LLVM::detail::initializeBackendsOnce()`
```

--------------------------------

### DenseElementsAttr::try_value_begin

Source: https://mlir.llvm.org/doxygen/mlir_2IR_2BuiltinAttributes_8h_source.html

Attempts to get an iterator of the specified type T to the start of the held element values.

```APIDOC
## auto mlir::DenseElementsAttr::try_value_begin() const

### Description
Try to get an iterator of the given type to the start of the held element values.

### Definition
BuiltinAttributes.h:393
```

--------------------------------

### mlir::MemRefDescriptor::bufferPtr

Source: https://mlir.llvm.org/doxygen/X86Dialect_8cpp_source.html

Builds IR for getting the start address of the buffer represented by this memref.

```APIDOC
## Value bufferPtr(OpBuilder &builder, Location loc, const LLVMTypeConverter &converter, MemRefType type)

### Description
Builds IR for getting the start address of the buffer represented by this memref: memref....

### Parameters
* **builder** (OpBuilder &) - The operation builder.
* **loc** (Location) - The location for the IR.
* **converter** (const LLVMTypeConverter &) - The LLVM type converter.
* **type** (MemRefType) - The MemRef type.
```

--------------------------------

### startEmit

Source: https://mlir.llvm.org/doxygen/classmlir_1_1sparse__tensor_1_1CodegenEnv.html

Starts the emission process with a given strategy.

```APIDOC
## void startEmit(SparseEmitStrategy emitStrategy)

### Description
Starts the emission process with a given strategy.

### Parameters
- **emitStrategy** (SparseEmitStrategy) - The emission strategy to use.
```

--------------------------------

### Get AOS MemRef

Source: https://mlir.llvm.org/doxygen/SparseTensorDescriptor_8h_source.html

Retrieves the MemRef for the coordinate of pointers (AoS COO) starting from the specified level.

```cpp
Value getAOSMemRef() const {
  const Level cooStart = rType.getAoSCOOStart();
  assert(cooStart < rType.getLvlRank());
  return getMemRefField(SparseTensorFieldKind::CrdMemRef, cooStart);
}
```

--------------------------------

### init

Source: https://mlir.llvm.org/doxygen/functions_i.html

Initializes various components, including dependence graphs, GPU module serialization, and transform dialect extensions.

```APIDOC
## init

### Description
Initializes various components such as dependence graphs, GPU module serialization, and transform dialect extensions.

### Signature
init() : mlir::affine::MemRefDependenceGraph, mlir::NVVM::SerializeGPUModuleBase, mlir::ROCDL::SerializeGPUModuleBase, mlir::transform::TransformDialectExtension< DerivedTy, ExtraDialects >, TuneExtension
```

--------------------------------

### mgpuLaunchKernel

Source: https://mlir.llvm.org/doxygen/VulkanRuntimeWrappers_8cpp.html

Launches a Vulkan compute kernel on the GPU.

```APIDOC
## mgpuLaunchKernel()

### Description
Launches a Vulkan compute kernel.

### Signature
```cpp
VULKAN_WRAPPER_SYMBOL_EXPORT void mgpuLaunchKernel(void *_vkKernel_, size_t _gridX_, size_t _gridY_, size_t _gridZ_, size_t _, size_t _, size_t _, size_t _, void *_vkRuntimeManager_, void **_params_, void **_, size_t _paramsCount_)
```

### Parameters
- **_vkKernel_** (void *) - Handle to the Vulkan kernel.
- **_gridX_** (size_t) - Grid dimension X.
- **_gridY_** (size_t) - Grid dimension Y.
- **_gridZ_** (size_t) - Grid dimension Z.
- **_vkRuntimeManager_** (void *) - Handle to the Vulkan runtime manager.
- **_params_** (void **) - Array of kernel parameters.
- **_paramsCount_** (size_t) - Number of kernel parameters.
```

--------------------------------

### Get Perfectly Nested Loops

Source: https://mlir.llvm.org/doxygen/LoopUtils_8h.html

Retrieves a perfectly nested sequence of loops starting from a given root.

```APIDOC
## getPerfectlyNestedLoops

### Description
Get perfectly nested sequence of loops starting at root of loop nest (the first op being another AffineFor, and the second op - a terminator).

### Signature
void mlir::affine::getPerfectlyNestedLoops (SmallVectorImpl< AffineForOp > &nestedLoops, AffineForOp root)

### Parameters
* **nestedLoops** (SmallVectorImpl< AffineForOp > &) - Output vector to store the nested loops.
* **root** (AffineForOp) - The root loop operation.
```

--------------------------------

### Get Slice SubMap from AffineMap

Source: https://mlir.llvm.org/doxygen/IR_2AffineMap_8cpp_source.html

Extracts a submap from an AffineMap by specifying a start index and length for the results.

```cpp
AffineMap AffineMap::getSliceMap(unsigned start, unsigned length) const {
  return AffineMap::get(getNumDims(), getNumSymbols(),
                        getResults().slice(start, length), getContext());
}
```

--------------------------------

### init

Source: https://mlir.llvm.org/python-bindings/_sources/autoapi/mlir/_mlir_libs/_mlirExecutionEngine/index.rst.txt

Initializes the execution engine.

```APIDOC
## init

### Description
Initializes the execution engine.
```

--------------------------------

### ShuffleOp Down Example

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_gpu_ops_gen/index.html

Illustrates the start of a 'down' shuffle operation, which involves an arithmetic constant for lane calculation.

```mlir
%cst1 = arith.constant 1 : i32
```

--------------------------------

### mlir::LLVM::detail::initializeBackendsOnce

Source: https://mlir.llvm.org/doxygen/TargetUtils_8cpp.html

Idempotent helper to register and initialize all backends that LLVM has been configured to support.

```APIDOC
## Function: initializeBackendsOnce

### Description
Idempotent helper to register and initialize all backends that LLVM has been configured to support.

### Signature
`void mlir::LLVM::detail::initializeBackendsOnce()`
```

--------------------------------

### Basic TypeBuilder Declaration

Source: https://mlir.llvm.org/docs/DefiningDialects/AttributesAndTypes

This is a basic `TypeBuilder` example that generates a standard `get` method declaration for `MyType`.

```tablegen
  let builders = [
    TypeBuilder<(ins "int":$intParam)>, 
  ];

```

--------------------------------

### ExecutionEngine::initialize

Source: https://mlir.llvm.org/doxygen/classmlir_1_1ExecutionEngine.html

Initializes the ExecutionEngine, preparing it for JIT compilation and execution.

```APIDOC
## ExecutionEngine::initialize

### Description
Initialize the ExecutionEngine.

### Signature
```cpp
void initialize()
```
```

--------------------------------

### initialize

Source: https://mlir.llvm.org/doxygen/classmlir_1_1ExecutionEngine-members.html

Initializes the execution engine.

```APIDOC
## initialize

### Description
Initializes the execution engine.

### Signature
void initialize()
```

--------------------------------

### Tensor Rank Operation Example

Source: https://mlir.llvm.org/docs/Dialects/TensorOps

Demonstrates how to get the rank of a tensor. Supports tensors with static or dynamic shapes.

```mlir
%0 = tensor.rank %arg0 : tensor<*xf32>
%1 = tensor.rank %arg1 : tensor<?x?xf32>
```

--------------------------------

### Setup Debug Configuration from CL Options

Source: https://mlir.llvm.org/doxygen/CLOptionsSetup_8cpp_source.html

Creates a DebugConfig object populated with settings derived from command-line options. This is the primary way to initialize the debugging configuration.

```cpp
static std::string diag(const llvm::Value &value)
**Definition** ModuleImport.cpp:57
mlir::tracing::InstallDebugHandler::Impl
**Definition** CLOptionsSetup.cpp:83
mlir::tracing::InstallDebugHandler::Impl::Impl
Impl(MLIRContext &context, const DebugConfig &config)
**Definition** CLOptionsSetup.cpp:85
mlir::MLIRContext
MLIRContext is the top-level object for a collection of MLIR operations.
**Definition** MLIRContext.h:63
mlir::MLIRContext::registerActionHandler
void registerActionHandler(HandlerTy handler)
Register a handler for handling actions that are dispatched through this context.
**Definition** MLIRContext.cpp:379
mlir::tracing::DebugConfig
**Definition** CLOptionsSetup.h:22
mlir::tracing::DebugConfig::getLogActionsLocFilters
ArrayRef< tracing::BreakpointManager * > getLogActionsLocFilters() const
Get the location breakpoint managers to use to filter out action logging.
**Definition** CLOptionsSetup.h:64
mlir::tracing::DebugConfig::createFromCLOptions
static DebugConfig createFromCLOptions()
Create a new config with the default set from the CL options.
**Definition** CLOptionsSetup.cpp:81
```

--------------------------------

### Get Array of Structs COO Start Level

Source: https://mlir.llvm.org/doxygen/SparseTensorDialect_8cpp_source.html

Retrieves the starting level index for an Array of Structs (AoS) COO segment. If no AoS segment is found or the COO segments are empty, it returns the total level rank.

```cpp
Level mlir::sparse_tensor::SparseTensorEncodingAttr::getAoSCOOStart() const {
SmallVector<COOSegment> coo = getCOOSegments();
assert(coo.size() == 1 || coo.empty());
if (!coo.empty() && coo.front().isAoS()) {
return coo.front().lvlRange.first;
}
return getLvlRank();
}
```

--------------------------------

### nvgpu.mbarrier.init

Source: https://mlir.llvm.org/docs/Dialects/NVGPU

Initializes an nvgpu.mbarrier object with a specified number of threads and an optional predicate.

```APIDOC
## nvgpu.mbarrier.init (nvgpu::MBarrierInitOp)

### Description
Initializes the `nvgpu.mbarrier` object with the given number of threads. An optional predicate can also be provided.

### Syntax
```
operation ::= `nvgpu.mbarrier.init` $barriers `[` $mbarId `]` `,` $count (`,` `predicate` `=` $predicate^)? attr-dict `:` type($barriers)
```

### Operands
#### Path Parameters
- **barriers** (mbarrier barrier type) - Description: mbarrier barrier type
- **mbarId** (index) - Description: index
- **count** (index) - Description: index
- **predicate** (1-bit signless integer) - Optional: 1-bit signless integer

### Example
```
  %num_threads = gpu.block_dim x
  %barrier = nvgpu.mbarrier.create -> !nvgpu.mbarrier.barrier<memorySpace = #gpu.address_space<workgroup>>
  nvgpu.mbarrier.init %barrier, %num_threads : !nvgpu.mbarrier.barrier<memorySpace = #gpu.address_space<workgroup>>
```
```

--------------------------------

### Get Number of Lines in a Chunk

Source: https://mlir.llvm.org/doxygen/PDLLServer_8cpp_source.html

Calculates the number of lines within a given chunk by subtracting the start line offset of the current chunk from the start line offset of the next chunk. This is useful for determining the extent of a chunk.

```cpp
auto getNumLines = [](ChunkIterator chunkIt) {
return (chunkIt + 1)->lineOffset - chunkIt->lineOffset;
};
```

--------------------------------

### Get Root Operation Position

Source: https://mlir.llvm.org/doxygen/lib_2Conversion_2PDLToPDLInterp_2Predicate_8h_source.html

Retrieves the root operation position. This is a fundamental starting point for defining predicate positions.

```cpp
Position *getRoot() { return OperationPosition::getRoot(uniquer); }
```

--------------------------------

### init()

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_acc_ops_gen/index.html

Helper function to create an InitOp.

```APIDOC
## init()

### Description
Helper function to create an InitOp.
```

--------------------------------

### mlir::MlirOptMainConfig::createFromCLOptions

Source: https://mlir.llvm.org/doxygen/MlirOptMain_8h_source.html

Creates a new MlirOptMainConfig instance initialized with default settings derived from command-line options.

```APIDOC
## createFromCLOptions

### Description
Creates a new configuration object populated with default settings parsed from command-line options.

### Signature
static MlirOptMainConfig createFromCLOptions()

### Definition
MlirOptMain.cpp:353
```

--------------------------------

### VPReduceOrOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector OR reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceOrOp

### Description
Performs a bitwise OR reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### Get Offset from MemRef

Source: https://mlir.llvm.org/doxygen/MemRefBuilder_8cpp_source.html

Retrieves the offset value from a MemRef descriptor. The offset indicates the starting position within the allocated buffer.

```cpp
Value UnrankedMemRefDescriptor::offset(OpBuilder &builder, Location loc,
 const LLVMTypeConverter &typeConverter,
 Value memRefDescPtr,
 LLVM::LLVMPointerType elemPtrType) {
 Value offsetPtr = 
 offsetBasePtr(builder, loc, typeConverter, memRefDescPtr, elemPtrType);
 return LLVM::LoadOp::create(builder, loc, typeConverter.getIndexType(),
 offsetPtr);
}
```

--------------------------------

### initialize

Source: https://mlir.llvm.org/doxygen/classmlir_1_1remark_1_1detail_1_1RemarkEngine-members.html

Initializes the RemarkEngine with a streamer, emitting policy, and an error message string.

```APIDOC
## initialize

### Description
Initializes the RemarkEngine with a streamer, emitting policy, and an error message string.

### Method
[Method Signature]
initialize(std::unique_ptr< MLIRRemarkStreamerBase > streamer, std::unique_ptr< RemarkEmittingPolicyBase > remarkEmittingPolicy, std::string *errMsg)

### Parameters
* **streamer** (std::unique_ptr< MLIRRemarkStreamerBase >) - The remark streamer.
* **remarkEmittingPolicy** (std::unique_ptr< RemarkEmittingPolicyBase >) - The remark emitting policy.
* **errMsg** (std::string *) - A string to store any error messages.
```

--------------------------------

### MlirOptMainConfig::createFromCLOptions

Source: https://mlir.llvm.org/doxygen/classmlir_1_1MlirOptMainConfig.html

Creates a new MlirOptMainConfig instance populated with options from the command line.

```APIDOC
## createFromCLOptions()

### Description
Creates a new configuration object by parsing options from the command line arguments.

### Method
`static MlirOptMainConfig createFromCLOptions()`

### Returns
A `MlirOptMainConfig` object initialized with command-line options.
```

--------------------------------

### Get Stride from MemRefDescriptorView

Source: https://mlir.llvm.org/doxygen/MemRefBuilder_8cpp_source.html

Retrieves a specific stride value from a MemRefDescriptorView at the given position. The position is relative to the start of the stride elements.

```C++
Value MemRefDescriptorView::stride(unsigned pos) {
  return elements[kSizePosInMemRefDescriptor + rank + pos];
}
```

--------------------------------

### mgpuLaunchKernel() Function

Source: https://mlir.llvm.org/doxygen/LevelZeroRuntimeWrappers_8cpp.html

Launches a Level Zero kernel on the device.

```APIDOC
## Function Documentation
## ◆ mgpuLaunchKernel()
void mgpuLaunchKernel(ze_kernel_handle_t kernel, size_t gridX, size_t gridY, size_t gridZ, size_t blockX, size_t blockY, size_t blockZ, size_t sharedMemBytes, StreamWrapper *stream, void **params, void **, size_t paramsCount)

**Description:**
This function launches a specified Level Zero kernel with given dimensions for the grid and thread blocks. It also allows configuration of shared memory, associates the launch with a stream, and passes kernel arguments. The `L0_SAFE_CALL` macro ensures safe execution.
```

--------------------------------

### Get Size from MemRefDescriptorView

Source: https://mlir.llvm.org/doxygen/MemRefBuilder_8cpp_source.html

Retrieves a specific size value from a MemRefDescriptorView at the given position. The position is relative to the start of the size elements.

```C++
Value MemRefDescriptorView::size(unsigned pos) {
  return elements[kSizePosInMemRefDescriptor + pos];
}
```

--------------------------------

### begin()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1SymbolTable_1_1UseRange.html

Returns an iterator to the beginning of the SymbolUse range.

```APIDOC
## begin()

| iterator mlir::SymbolTable::UseRange::begin | ( | __ | ) |  const  
---|---|---|---|---
inline  
Definition at line 208 of file SymbolTable.h.
```

--------------------------------

### Get Top-Level Tileable Bands

Source: https://mlir.llvm.org/doxygen/LoopTiling_8cpp_source.html

Retrieves maximal perfect nests of 'affine.for' operations starting from the root of a function that are valid for tiling.

```cpp
static void
getTopLevelTileableBands(func::FuncOp f,
 std::vector<SmallVector<AffineForOp, 6>> &bands) {
 // Get maximal perfect nest of 'affine.for' ops starting from root
 // (inclusive).
 for (AffineForOp forOp : f.getOps<AffineForOp>()) {
 SmallVector<AffineForOp, 6> band;
 getPerfectlyNestedLoops(band, forOp);
 if (isTilingValid(band))
 bands.push_back(band);
 }
}
```

--------------------------------

### VPReduceMulOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector multiply reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceMulOp

### Description
Performs a multiply reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### MLIR AffineMap Get

Source: https://mlir.llvm.org/doxygen/LowerVectorContract_8cpp_source.html

Static method to create a basic AffineMap with no dimensions or symbols. This is a foundational map often used as a starting point.

```cpp
static AffineMap get(MLIRContext *context)
```

--------------------------------

### Initialize All MLIR Extensions

Source: https://mlir.llvm.org/doxygen/CAPI_2RegisterEverything_2RegisterEverything_8cpp.html

Include the header for initializing all MLIR extensions. Extensions provide additional features and capabilities to MLIR.

```c
#include "mlir/InitAllExtensions.h"
```

--------------------------------

### MlirOptMain (Argc, Argv, InputFilename, OutputFilename, Registry)

Source: https://mlir.llvm.org/doxygen/MlirOptMain_8h_source.html

An implementation for `mlir-opt` that allows CLI options to be accessed before running MlirOptMain, using input and output filenames.

```APIDOC
## MlirOptMain (Argc, Argv, InputFilename, OutputFilename, Registry)

### Description
Implementation for tools like `mlir-opt`. This function can be used with registerAndParseCLIOptions so that CLI options can be accessed before running MlirOptMain. - inputFilename is the name of the input mlir file. - outputFilename is the name of the output file.

### Parameters
- **argc** (int) - The argument count.
- **argv** (char **) - The argument values.
- **inputFilename** (llvm::StringRef) - The name of the input MLIR file.
- **outputFilename** (llvm::StringRef) - The name of the output file.
- **registry** (DialectRegistry &) - The dialect registry.

### Return Value
LogicalResult - Success if the processing is successful, failure otherwise.
```

--------------------------------

### VPReduceUMinOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector unsigned minimum reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceUMinOp

### Description
Performs an unsigned minimum reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### VPReduceUMaxOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector unsigned maximum reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceUMaxOp

### Description
Performs an unsigned maximum reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### init()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1NVVM_1_1SerializeGPUModuleBase-members.html

Static method to initialize the serialization process.

```APIDOC
## init()

### Description
Static method to initialize the serialization process.

### Method
static void init()

### Parameters
None

### Returns
None
```

--------------------------------

### VPReduceSMinOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector signed minimum reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceSMinOp

### Description
Performs a signed minimum reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### createGpuLaunch

Source: https://mlir.llvm.org/doxygen/namespacemlir_1_1transform_1_1gpu.html

Creates an empty-body gpu::LaunchOp using the provided kernel settings and places a terminator within.

```APIDOC
## createGpuLaunch

### Description
Create an empty-body gpu::LaunchOp using the provided kernel settings and put a terminator within.

### Signature
```cpp 
DiagnosedSilenceableFailure createGpuLaunch(RewriterBase &rewriter, Location loc, TransformOpInterface transformOp, mlir::gpu::LaunchOp &launchOp, std::optional< int64_t > gridDimX = std::nullopt, std::optional< int64_t > gridDimY = std::nullopt, std::optional< int64_t > gridDimZ = std::nullopt, std::optional< int64_t > blockDimX = std::nullopt, std::optional< int64_t > blockDimY = std::nullopt, std::optional< int64_t > blockDimZ = std::nullopt)
```
```

```APIDOC
## createGpuLaunch (Overload)

### Description
Create an empty-body gpu::LaunchOp using the provided kernel settings and put a terminator within.

### Signature
```cpp 
DiagnosedSilenceableFailure createGpuLaunch(RewriterBase &rewriter, Location loc, TransformOpInterface transformOp, LaunchOp &launchOp, std::optional< int64_t > gridDimX, std::optional< int64_t > gridDimY, std::optional< int64_t > gridDimZ, std::optional< int64_t > blockDimX, std::optional< int64_t > blockDimY, std::optional< int64_t > blockDimZ)
```
```

--------------------------------

### nvgpu.mbarrier.init

Source: https://mlir.llvm.org/python-bindings/_sources/autoapi/mlir/dialects/nvgpu/index.rst.txt

Initializes an mbarrier object. This operation is used to set up a new mbarrier for synchronization.

```APIDOC
## nvgpu.mbarrier.init

### Description
Initializes an mbarrier object.

### Syntax
```mlir
"nvgpu.mbarrier.init"(%barriers, %count, %mbar_id) : (!nvgpu.mbarrier.barrier, _I64, _I64) -> !nvgpu.mbarrier.init
```

### Operands
- **barriers**: The mbarrier object to initialize.
- **count**: The number of threads to synchronize.
- **mbar_id**: The ID of the mbarrier.

### Attributes
None
```

--------------------------------

### VPReduceSMaxOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector signed maximum reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceSMaxOp

### Description
Performs a signed maximum reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### printLaunchFuncOperands

Source: https://mlir.llvm.org/doxygen/GPUDialect_8cpp_source.html

Prints the 'args' keyword and the list of kernel function operands and their types in assembly format.

```APIDOC
## printLaunchFuncOperands

### Description
Prints the 'args' keyword followed by a parenthesized list of kernel function operands and their types in the assembly format. If no operands are present, nothing is printed.

### Signature
`static void printLaunchFuncOperands(OpAsmPrinter &printer, Operation *op, OperandRange operands, TypeRange types)`
```

--------------------------------

### VPReduceFMulOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector floating-point multiply reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceFMulOp

### Description
Performs a floating-point multiply reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### LaunchFuncOp::build (GPUFuncOp overload)

Source: https://mlir.llvm.org/doxygen/GPUDialect_8cpp_source.html

Builds a `LaunchFuncOp` using a `GPUFuncOp` directly.

```APIDOC
## LaunchFuncOp::build

### Description
Constructs a `LaunchFuncOp` instance. This overload takes a `GPUFuncOp` directly, along with grid and block dimensions, dynamic shared memory size, kernel operands, and an optional cluster size. It resolves the `GPUFuncOp` to a `SymbolRefAttr`.

### Parameters
- `builder` (OpBuilder &): The MLIR operation builder.
- `result` (OperationState &): The state of the operation being built.
- `kernelFunc` (GPUFuncOp): The GPU function to launch.
- `gridSize` (KernelDim3): The grid dimensions (x, y, z).
- `getBlockSize` (KernelDim3): The block dimensions (x, y, z).
- `dynamicSharedMemorySize` (Value): The value representing the dynamic shared memory size.
- `kernelOperands` (ValueRange): The operands for the kernel.
- `asyncTokenType` (Type): The type of the async token, if applicable.
- `asyncDependencies` (ValueRange): Dependencies for the asynchronous operation.
- `clusterSize` (std::optional<KernelDim3>): Optional cluster dimensions (x, y, z).
```

--------------------------------

### VPReduceFMinOp

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_llvm_ops_gen/index.html

Represents the vector floating-point minimum reduction operation. It supports operations like getting the start value, the value, mask, evaluation, and result.

```APIDOC
## VPReduceFMinOp

### Description
Performs a floating-point minimum reduction on a vector.

### Attributes
- `start_value`: The initial value for the reduction.
- `val`: The input vector for the reduction.
- `mask`: The mask to apply to the reduction.
- `evl`: The evaluation mode.
- `res`: The result of the reduction.
```

--------------------------------

### Get Root Timer - TimingManager

Source: https://mlir.llvm.org/doxygen/Timing_8h_source.html

Retrieves the root timer from the TimingManager. This timer must be manually started and stopped. Use getRootScope() for RAII-style management.

```cpp
Timer getRootTimer();
```

--------------------------------

### startExpand()

Source: https://mlir.llvm.org/doxygen/classmlir_1_1sparse__tensor_1_1CodegenEnv.html

Initiates the expansion process with provided values for filled, added, and count.

```APIDOC
## startExpand()

### Description
Initiates the expansion process with provided values for filled, added, and count.

### Method
`void CodegenEnv::startExpand(Value _values_, Value _filled_, Value _added_, Value _count_)`

### Definition
`lib/Dialect/SparseTensor/Transforms/Utils/CodegenEnv.cpp:204`
```

--------------------------------

### _site_initialize()

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/_mlir_libs/index.html

Initializes the site-specific MLIR components.

```APIDOC
## _site_initialize()

### Description
Initializes the site-specific MLIR components.
```

--------------------------------

### Get Cycle in Parent Relation

Source: https://mlir.llvm.org/doxygen/RootOrdering_8cpp_source.html

Retrieves the cycle formed by a parent relation starting from a given node. Asserts that the cycle does not contain empty values.

```cpp
static SmallVector<Value> getCycle(const DenseMap<Value, Value> &parents,
 Value rep) {
 SmallVector<Value> cycle;
 Value node = rep;
 do {
 cycle.push_back(node);
 node = parents.lookup(node);
 assert(node && "got an empty value in the cycle");
 } while (node != rep);
 return cycle;
}
```

--------------------------------

### Memref DimOp Example

Source: https://mlir.llvm.org/python-bindings/autoapi/mlir/dialects/_memref_ops_gen/index.html

Demonstrates the usage of the memref.dim operation to get the size of a dimension of a memref. This operation is useful for dynamic memory access and manipulation.

```mlir
memref.dim %memref, %index
```

--------------------------------

### initialize

Source: https://mlir.llvm.org/doxygen/classConvertSameRankInsertStridedSliceIntoShuffle-members.html

Initializes the pattern.

```APIDOC
## initialize

### Description
Initializes the pattern.

### Method
inline

### Signature
initialize()
```

--------------------------------

### RemarkEngine::initialize

Source: https://mlir.llvm.org/doxygen/Remarks_8cpp_source.html

Sets up the remark engine with a specified output streamer and remark emitting policy. This is a crucial step before any remarks can be reported.

```APIDOC
## initialize

### Description
Setup the remark engine with the given output path and format.

### Method
`LogicalResult initialize(std::unique_ptr< MLIRRemarkStreamerBase > streamer, std::unique_ptr< RemarkEmittingPolicyBase > remarkEmittingPolicy, std::string *errMsg)`

### Parameters
* `streamer` (std::unique_ptr< MLIRRemarkStreamerBase >) - The streamer to use for outputting remarks.
* `remarkEmittingPolicy` (std::unique_ptr< RemarkEmittingPolicyBase >) - The policy to control remark emission.
* `errMsg` (std::string *) - A pointer to a string for error messages.
```

--------------------------------

### Get AffineMap Slice

Source: https://mlir.llvm.org/doxygen/mlir_2IR_2AffineMap_8h_source.html

Returns a new AffineMap consisting of a specified number of results, starting from a given index. This allows extracting sub-maps from larger ones.

```cpp
AffineMap getSliceMap(unsigned start, unsigned length) const

```

--------------------------------

### createGpuLaunch

Source: https://mlir.llvm.org/doxygen/GPUTransformOps_8cpp_source.html

Creates an empty-body gpu::LaunchOp using the provided kernel settings and inserts a terminator.

```APIDOC
## createGpuLaunch

### Description
Creates an empty-body gpu::LaunchOp using the provided kernel settings and put a terminator within.

### Signature
DiagnosedSilenceableFailure createGpuLaunch(RewriterBase &rewriter, Location loc, TransformOpInterface transformOp, mlir::gpu::LaunchOp &launchOp, std::optional< int64_t > gridDimX=std::nullopt, std::optional< int64_t > gridDimY=std::nullopt, std::optional< int64_t > gridDimZ=std::nullopt, std::optional< int64_t > blockDimX=std::nullopt, std::optional< int64_t > blockDimY=std::nullopt, std::optional< int64_t > blockDimZ=std::nullopt)
```

--------------------------------

### Get Collapsed Indices for Shape Collapse

Source: https://mlir.llvm.org/doxygen/VectorTransferOpTransforms_8cpp_source.html

Returns the new indices that collapse the inner dimensions of a shape starting from a specified dimension. This is a helper for collapse_shape operations.

```cpp
static SmallVector< Value > getCollapsedIndices(RewriterBase &rewriter, Location loc, ArrayRef< int64_t > shape, ValueRange indices, int64_t firstDimToCollapse)
```

--------------------------------

### Get Symbol Uses Implementation

Source: https://mlir.llvm.org/doxygen/SymbolTable_8cpp_source.html

Internal implementation for retrieving symbol uses. It walks symbol uses starting from a given operation or region and collects them into a vector.

```cpp
template <typename FromT>
static std::optional<SymbolTable::UseRange> getSymbolUsesImpl(FromT from) {
  std::vector<SymbolTable::SymbolUse> uses;
  auto walkFn = [&](SymbolTable::SymbolUse symbolUse) {
    uses.push_back(symbolUse);
    return WalkResult::advance();
  };
  auto result = walkSymbolUses(from, walkFn);
  return result ? std::optional<SymbolTable::UseRange>(std::move(uses))
                 : std::nullopt;
}
```

--------------------------------

### Remark System Setup

Source: https://mlir.llvm.org/doxygen/Remarks_8h.html

Function to enable and configure the optimization remark system for an MLIR context.

```APIDOC
## mlir::remark::enableOptimizationRemarks

### Description
Setup remarks for the context. This function initializes the remarking system, allowing for the emission of optimization remarks.

### Signature
```cpp
LogicalResult mlir::remark::enableOptimizationRemarks(MLIRContext &ctx, std::unique_ptr< remark::detail::MLIRRemarkStreamerBase > streamer, std::unique_ptr< remark::detail::RemarkEmittingPolicyBase > remarkEmittingPolicy, const remark::RemarkCategories &cats, bool printAsEmitRemarks = false)
```

### Parameters

*   **ctx** (MLIRContext &) - The MLIR context to associate the remark system with.
*   **streamer** (std::unique_ptr< remark::detail::MLIRRemarkStreamerBase >) - The underlying remark streamer to use.
*   **remarkEmittingPolicy** (std::unique_ptr< remark::detail::RemarkEmittingPolicyBase >) - The policy that determines which remarks to emit.
*   **cats** (const remark::RemarkCategories &) - The set of remark categories to accept.
*   **printAsEmitRemarks** (bool) - Whether to print remarks as if they were emitted.
```