TTMetalOp - tt-mlir documentation

`ttmetal.create_buffer` (tt::ttmetal::CreateBufferOp)

Create buffer op.

Create buffer operation.

When this buffer uses a virtual grid, virtualGridInverseMapping stores the inverse affine map (physical to virtual grid coordinates) and virtualGridForwardMapping stores the forward affine map (virtual to physical grid coordinates). Both are propagated from d2m.empty through bufferization.

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Allocate on ::mlir::SideEffects::DefaultResource}

Attributes:

Attribute	MLIR Type	Description
`address`	::mlir::IntegerAttr	64-bit signless integer attribute
`virtualGridInverseMapping`	::mlir::AffineMapAttr	AffineMap attribute
`virtualGridForwardMapping`	::mlir::AffineMapAttr	AffineMap attribute

Results:

Result	Description
`result`	non-0-ranked.memref of any type values

`ttmetal.create_global_semaphore` (tt::ttmetal::CreateGlobalSemaphoreOp)

Create global semaphore op.

Create global semaphore operation

Interfaces: InferTypeOpInterface, MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

Attributes:

Attribute	MLIR Type	Description
`address`	::mlir::IntegerAttr	64-bit signless integer attribute
`initial_value`	::mlir::IntegerAttr	32-bit unsigned integer attribute
`core_range`	::mlir::tt::ttmetal::CoreRangeAttr	TTMetal grid attribute {{% markdown %}} TTMetal grid attribute {{% /markdown %}}

Results:

Result	Description
`result`	TTMetal global semaphore

`ttmetal.deallocate_buffer` (tt::ttmetal::DeallocateBufferOp)

Deallocate buffer op.

Deallocate buffer operation

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Free on ::mlir::SideEffects::DefaultResource}

Operands:

Operand	Description
`input`	non-0-ranked.memref of any type values

`ttmetal.enqueue_program` (tt::ttmetal::EnqueueProgramOp)

Enqueue program op.

Enqueue program operation

Traits: AttrSizedOperandSegments

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Read on ::mlir::SideEffects::DefaultResource, MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

Attributes:

Attribute	MLIR Type	Description
`cb_ports`	::mlir::DenseI64ArrayAttr	i64 dense array attribute
`kernelConfigs`	::mlir::ArrayAttr
`fabricConnectionConfig`	::mlir::tt::ttmetal::FabricConnectionConfigAttr	TTMetal CCL Config attribute. {{% markdown %}} Structure for configuring fabric connections for worker cores (in CCL ops). {{% /markdown %}}

Operands:

Operand	Description
`args`	variadic of any type
`cbs`	variadic of non-0-ranked.memref of any type values

`ttmetal.enqueue_read_buffer` (tt::ttmetal::EnqueueReadBufferOp)

Enqueue read buffer op.

Enqueue read buffer operation

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Read on ::mlir::SideEffects::DefaultResource, MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

Operands:

Operand	Description
`input`	non-0-ranked.memref of any type values
`output`	non-0-ranked.memref of any type values

`ttmetal.enqueue_write_buffer` (tt::ttmetal::EnqueueWriteBufferOp)

Enqueue write buffer op.

Enqueue write buffer operation

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Read on ::mlir::SideEffects::DefaultResource, MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

Operands:

Operand	Description
`input`	non-0-ranked.memref of any type values
`output`	non-0-ranked.memref of any type values

`ttmetal.finish` (tt::ttmetal::FinishOp)

Finish op for command queue.

Global barrier op, used to wait for all commands on queue to finish.

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Read on ::mlir::SideEffects::DefaultResource, MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

`ttmetal.mesh_shard` (tt::ttmetal::MeshShardOp)

Nd sharding or (partial) concat op

Nd sharding or (partial) concat op in D2M runtime. ShardToFull: Nd sharding in host memory. FullToshard: (partial) concat in host memory.

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Read on ::mlir::SideEffects::DefaultResource, MemoryEffects::Allocate on ::mlir::SideEffects::DefaultResource}

Attributes:

Attribute	MLIR Type	Description
`shard_type`	::mlir::tt::ttcore::MeshShardTypeAttr	MeshShard shard_type attribute in TT dialect {{% markdown %}} Define sharded tensor data of mesh_shard op. - Identity: input and output tensors are pre-sharded (same data) and no sharding is required. - Replicate: all of the devices has full tensor (same data). - Maximal: one or part of the devcices has full tensor (same data). - Devices: all or part of the devices has sharded (partial) tensor (different data). {{% /markdown %}}
`shard_direction`	::mlir::tt::ttcore::MeshShardDirectionAttr	TT MeshShardDirection
`shard_shape`	::mlir::DenseI64ArrayAttr	i64 dense array attribute
`shard_dims`	::mlir::DenseI64ArrayAttr	i64 dense array attribute

Operands:

Operand	Description
`input`	non-0-ranked.memref of any type values

Results:

Result	Description
`result`	non-0-ranked.memref of any type values

`ttmetal.reset_global_semaphore` (tt::ttmetal::ResetGlobalSemaphoreOp)

Reset global semaphore op.

Reset global semaphore operation

Interfaces: MemoryEffectOpInterface (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{MemoryEffects::Write on ::mlir::SideEffects::DefaultResource}

Attributes:

Attribute	MLIR Type	Description
`value`	::mlir::IntegerAttr	32-bit unsigned integer attribute

Operands:

Operand	Description
`semaphore`	TTMetal global semaphore