'mhlo' Dialect

Stay organized with collections Save and categorize content based on your preferences.

Operation definition

mhlo.abs (::mlir::mhlo::AbsOp)

Absolute value operator

Returns abs(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32/64-bit signless integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 4/8/16/32/64-bit signless integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.add_dependency (::mlir::mhlo::AddDependencyOp)

AddDependency operator

Syntax:

operation ::= `mhlo.add_dependency` operands attr-dict `:` functional-type(operands, results)

AddDependency takes two operands: a data operand and a token. The output of the operation is the data operand. When used with AfterAll this operation enables ordering non-side-effecting operations (those that do not produce token values).

Example:

%1 = mhlo.add_dependency %arg0, %0 : (tensor<3x4xf32>, !mhlo.token) -> tensor<3x4xf32>

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token
token token

Results:

Result Description
output tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.add (::mlir::mhlo::AddOp)

Addition operator

Returns lhs + rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.after_all (::mlir::mhlo::AfterAllOp)

AfterAll operator

Syntax:

operation ::= `mhlo.after_all` operands attr-dict `:` functional-type(operands, results)

AfterAll takes a variadic number of tokens and produces a single token. Tokens are primitive types which can be threaded between side-effecting operations to enforce ordering. AfterAll can be used as a join of tokens for ordering a operation after a set operations.

See https://www.tensorflow.org/xla/operation_semantics#afterall

Example:

%0 = mhlo.after_all %arg0, %arg1 : (!mhlo.token, !mhlo.token) -> !mhlo.token

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operands token

Results:

Result Description
«unnamed» token

mhlo.all_gather (::mlir::mhlo::AllGatherOp)

AllGather operator

Performs concatenation across replicas.

See https://www.tensorflow.org/xla/operation_semantics#allgather

Traits: SameOperandsAndResultElementType

Attributes:

Attribute MLIR Type Description
all_gather_dim ::mlir::IntegerAttr 64-bit signless integer attribute
replica_groups ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
channel_handle ::mlir::mhlo::ChannelHandleAttr two 64-bit integers 'handle' and 'type'
use_global_device_ids ::mlir::UnitAttr unit attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.all_reduce (::mlir::mhlo::AllReduceOp)

AllReduce operator

Performs a custom reduction across replicas.

See https://www.tensorflow.org/xla/operation_semantics#allreduce

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
replica_groups ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
channel_handle ::mlir::mhlo::ChannelHandleAttr two 64-bit integers 'handle' and 'type'
use_global_device_ids ::mlir::UnitAttr unit attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.all_to_all (::mlir::mhlo::AllToAllOp)

Traits: InferTensorType, SameOperandsElementType, SameOperandsShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
split_dimension ::mlir::IntegerAttr 64-bit signless integer attribute
concat_dimension ::mlir::IntegerAttr 64-bit signless integer attribute
split_count ::mlir::IntegerAttr 64-bit signless integer attribute
replica_groups ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.and (::mlir::mhlo::AndOp)

And operator

Returns biwise-AND of lhs and rhs element-wise. The input tensors must be of type integer HLO_Int or boolean HLO_Pred.

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
rhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.async_done (::mlir::mhlo::AsyncDoneOp)

AsyncDone operator

AsyncDone blocks until the end of an asynchronous computation. It returns the final result of the asynchronous computation.

See the documentation for AsyncStart for more information.

Interfaces: InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
called_computation ::mlir::FlatSymbolRefAttr flat symbol reference attribute
execution_thread ::mlir::StringAttr string attribute
group_id ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
bundle async_bundle with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.async_start (::mlir::mhlo::AsyncStartOp)

AsyncStart operator

AsyncStart kicks off an asynchronous computation.

This is used when there are functions that contain both asynchronous waits (such as DMAs) and on-thread computation. For example, a function might consist of a computation, a DMA, another computation, a second DMA, and a final computation. This would be represented as an async_start followed by and async_update and an async_done. The async_start would do the first computation on-thread and then start the DMA. The async_update would wait for the DMA to complete if it wasn't yet done, then execute the second computation in the function, and start the second DMA. Finally, the async_done would wait on this last DMA, and then run the last computation that needs to be run on-thread and return the result of that final computation.

operands are passed to the computation directly called_computation is the function that will be run asynchronously execution_thread is the name of the thread in which it will be run. The main thread is called "main". All threads have names. group_id labels a set of async-start, async-done, and zero or more async-update ops corresponding to the same computation. We represent a missing group_id with either an negative value or None.

This returns all the state needed between async ops. After buffer assignment, the return values represents the space needed to hold the input, results, and any scratchpads needed or edited by the async op.

Attributes:

Attribute MLIR Type Description
called_computation ::mlir::FlatSymbolRefAttr flat symbol reference attribute
execution_thread ::mlir::StringAttr string attribute
group_id ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» async_bundle with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.async_update (::mlir::mhlo::AsyncUpdateOp)

AsyncUpdate operator

AsyncUpdate blocks on an asynchronous computation until a sync barrier. This returns bundle after operating on it.

See the documentation for AsyncStart for more information.

Interfaces: InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
called_computation ::mlir::FlatSymbolRefAttr flat symbol reference attribute
execution_thread ::mlir::StringAttr string attribute
group_id ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
bundle async_bundle with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» async_bundle with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.atan2 (::mlir::mhlo::Atan2Op)

Atan2 operator

Returns atan2(lhs/rhs) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.batch_norm_grad (::mlir::mhlo::BatchNormGradOp)

Batch Normalization Gradient

Calculates gradients of batch norm.

See https://www.tensorflow.org/xla/operation_semantics#batchnormgrad

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
epsilon ::mlir::FloatAttr 32-bit float attribute
feature_index ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
scale 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
mean 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
variance 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
grad_output ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
grad_operand ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
grad_scale 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
grad_offset 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.batch_norm_inference (::mlir::mhlo::BatchNormInferenceOp)

Batch Normalization for Inference

Normalizes an array across batch and spatial dimensions.

See https://www.tensorflow.org/xla/operation_semantics#batchnorminference

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
epsilon ::mlir::FloatAttr 32-bit float attribute
feature_index ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
scale 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
offset 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
mean 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
variance 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.batch_norm_training (::mlir::mhlo::BatchNormTrainingOp)

Batch Normalization for Training

Normalizes an array across batch and spatial dimensions.

See https://www.tensorflow.org/xla/operation_semantics#batchnormtraining

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
epsilon ::mlir::FloatAttr 32-bit float attribute
feature_index ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
scale 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
offset 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
output ranked tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
batch_mean 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
batch_var 1D tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.bitcast_convert (::mlir::mhlo::BitcastConvertOp)

BitcastConvert operator

Syntax:

operation ::= `mhlo.bitcast_convert` operands attr-dict `:` functional-type(operands, results)

Similar to a 'tf.bitcast' in TensorFlow, performs an element-wise bitcast operation from a data shape to a target shape. The dimensions must match, and the conversion is an element-wise one. Bitcast is implemented as a low-level cast, so machines with different floating-point representations will give different results.

See https://www.tensorflow.org/xla/operation_semantics#bitcastconverttype

Example:

%0 = mhlo.bitcast_convert %arg0 : (tensor<2xi32>) -> tensor<2xf32>

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.bitcast (::mlir::mhlo::BitcastOp)

Bitcast operator

Syntax:

operation ::= `mhlo.bitcast` operands attr-dict `:` functional-type(operands, results)

This op changes the shape of the input in the way that the physical arrangement of elements are unchanged.

However, the op needs layout information to make sense of "physical arrangement of elements". Layout support in MHLO is currently under exploration.

Example:

%0 = mhlo.bitcast %arg0 : (tensor<3x4xf32>) -> tensor<3x4x1xf32>

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.broadcast_in_dim (::mlir::mhlo::BroadcastInDimOp)

Broadcast a tensor into the given shape by adding dimensions.

Broadcasts the operand tensor to a higher rank. This is not the limited form of broadcasting exposed as the XLA client broadcast op, but rather the more powerful "InDim" broadcasting, which is closer to the HLO broadcast op and exposed in the XLA client BroadcastInDim method.

broadcast_dimensions maps the operand dimension number to the target shape dimension number. It must have the same size as the rank of the operand. The mapped dimensions must either be the same size or the dimension being broadcast from must be size 1 (degenerate broadcasting).

For a scalar (0D tensor) operand, broadcast_dimensions must be empty. The The scalar value will be broadcast to every element in the target shape.

See https://www.tensorflow.org/xla/broadcasting

Traits: SameOperandsAndResultElementType

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
broadcast_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» statically shaped tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.broadcast (::mlir::mhlo::BroadcastOp)

Broadcast a tensor to a higher rank by prepending dimensions

Broadcasts the operand tensor to a higher rank by prepending broadcast_sizes to the dimensions. The current values of the operand are copied into the other dimensions.

This is a more limited form of broadcasting, that corresponds to the XLA client Broadcast method. For a more general form of broadcasting, see the BroadcastInDimOp.

See https://www.tensorflow.org/xla/operation_semantics#broadcast

Traits: InferTensorType, SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
broadcast_sizes ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.case (::mlir::mhlo::CaseOp)

Switch-Case operator

Returns the result of executing branches[index]. If index is < 0 or >= N, then branches[N-1] is executed as the default branch.

The type of the returned values of each branch must be the same and equal to the types of the values returned by the operation.

Note that only one of the branches will be executed depending on the value of index.

Traits: RecursiveSideEffects, SingleBlockImplicitTerminator

Operands:

Operand Description
index tensor of 32-bit signless integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.cbrt (::mlir::mhlo::CbrtOp)

Cubic root operator

Returns element-wise cubic root of the operand.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.ceil (::mlir::mhlo::CeilOp)

Ceil operator

Returns Ceil(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.cholesky (::mlir::mhlo::CholeskyOp)

Cholesky operator

Computes the Cholesky decomposition of a batch of symmetric (Hermitian) positive definite matrices.

If lower is true, computes lower-triangular matrices l such that a=l.Transpose(l). If lower is false, computes upper-triangular matrices u such that a=Transpose(u).u.

Input data is read only from the lower/upper triangle of a, depending on the value of lower. Values from the other triangle are ignored. Output data is returned in the same triangle; the values in the other triangle are implementation-defined and may be anything.

If the rank of a is greater than 2, a is treated as a batch of matrices, where all except the minor 2 dimensions are batch dimensions.

If a is not symmetric (Hermitian) positive definite, the result is implementation-defined.

See https://www.tensorflow.org/xla/operation_semantics#cholesky

Traits: InferTensorType, SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
lower ::mlir::BoolAttr bool attribute

Operands:

Operand Description
a tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.clamp (::mlir::mhlo::ClampOp)

Clamp operator

Syntax:

operation ::= `mhlo.clamp` operands attr-dict `:` functional-type(operands, results)

Clamps an operand to within the range between a minimum and maximum value.

See https://www.tensorflow.org/xla/operation_semantics#clamp

Example:

%0 = mhlo.clamp %arg0, %arg1, %arg2 : (tensor<f32>, tensor<4xf32>, tensor<f32>) -> tensor<4xf32>

Traits: HLO_BroadcastingElementwise, InferTensorType, SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
min tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
max tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.count_leading_zeros (::mlir::mhlo::ClzOp)

Count-leading-zeros (Clz) operator

Returns the number of leading zeros in each operand element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

mhlo.collective_permute (::mlir::mhlo::CollectivePermuteOp)

CollectivePermute operator

CollectivePermute is a collective operation that sends and receives data cross replicas. Note that there are the following restrictions on the source_target_pair:

  • Any two pairs should not have the same target replica id, and they should not have the same source replica id.
  • If a replica id is not a target in any pair, then the output on that replica is a tensor consists of 0(s) with the same shape as the input.

See https://www.tensorflow.org/xla/operation_semantics#collectivepermute

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
source_target_pairs ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.compare (::mlir::mhlo::CompareOp)

Comparison operator

Syntax:

operation ::= `mhlo.compare` $comparison_direction `,` $lhs `,` $rhs (`,` $compare_type^)?
              attr-dict `:` functional-type(operands, results)

Compares lhs and rhs elementwise according to comparison_direction and compare_type. If unspecified, compare_type is FLOAT for float element types, SIGNED for signed element types and UNSIGNED for unsigned element types.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_comparison_operations

Example:

%0 = mhlo.compare LT, %arg0, %arg1 : (tensor<2xi32>, tensor<2xi32>) -> tensor<2xi1>
%1 = mhlo.compare LT, %arg0, %arg1, TOTALORDER : (tensor<2xi32>, tensor<2xi32>) -> tensor<2xi1>

Traits: Elementwise, InferTensorType, SameOperandsAndResultShape, SameOperandsElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
comparison_direction ::mlir::mhlo::ComparisonDirectionAttr Which comparison operation to perform.
compare_type ::mlir::mhlo::ComparisonTypeAttr Which comparison type to use.

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of pred (AKA boolean or 1-bit integer) values

mhlo.complex (::mlir::mhlo::ComplexOp)

Complex operator

Syntax:

operation ::= `mhlo.complex` `(`operands`)` attr-dict `:` `(`type(operands)`)` `->` type($result)

Performs element-wise conversion of a pair of real and imaginary values to a complex value.

Traits: Elementwise, SameOperandsAndResultShape, SameOperandsElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 32-bit float or 64-bit float values
rhs tensor of 32-bit float or 64-bit float values

Results:

Result Description
result tensor of complex type with 32-bit float or 64-bit float elements values

mhlo.compute_reshape_shape (::mlir::mhlo::ComputeReshapeShapeOp)

Compute input for reshape with any dynamic dim resolved

Syntax:

operation ::= `mhlo.compute_reshape_shape` $num_elements `,` $dynamic_shape attr-dict `:` type($num_elements) `,` type($dynamic_shape) `->` type($result)

This operation handles the dynamic aspect of a TF/NumPy/CHLO reshape. The dynamic aspect is that a single extent can be -1 and that dimension will instead be computed. This handles the computation and can then be passed to an HLO DynamicReshapeOp to replicate the TF/NumPy reshape behavior.

This op has undefined behavior if the dimensions do not evenly divide the number of elements, or if there are multiple -1 values. It is an identity op if no dimensions are -1.

%0 = hlo.compute_reshape_shape 12, [2, -1] -> [2, 6]

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
num_elements index
dynamic_shape 1D tensor of integer or index values

Results:

Result Description
result 1D tensor of integer or index values

mhlo.concatenate (::mlir::mhlo::ConcatenateOp)

XLA's concatenate op

Concatenates a set of tensors along the specified dimension.

See https://www.tensorflow.org/xla/operation_semantics#concatenate

Traits: SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimension ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
val tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.constant (::mlir::mhlo::ConstantOp)

Constant operator

Represents a constant value.

Traits: ConstantLike

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
value ::mlir::ElementsAttr constant vector/tensor attribute

Results:

Result Description
output statically shaped tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.convert (::mlir::mhlo::ConvertOp)

Convert operator

Performs element-wise conversion of values from one type to another, e.g. float to int.

See https://www.tensorflow.org/xla/operation_semantics#convertelementtype

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.convolution (::mlir::mhlo::ConvolutionOp)

Convolution operator

Syntax:

operation ::= `mhlo.convolution` `(`operands`)`
              `dim_numbers` `=` custom<ConvolutionDimensions>($dimension_numbers) `,`
              `window` `=` `{` custom<WindowAttributes>($window_strides, $padding,
              $lhs_dilation, $rhs_dilation,
              $window_reversal) `}`
              attr-dict `:` functional-type(operands, results)

Computes a convolution of the kind used in neural networks.

See https://www.tensorflow.org/xla/operation_semantics#conv_convolution

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
window_strides ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
padding ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
lhs_dilation ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
rhs_dilation ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
window_reversal ::mlir::DenseElementsAttr constant boolean vector/tensor attribute
dimension_numbers ::mlir::mhlo::ConvDimensionNumbersAttr Structure of dimension information for conv op
feature_group_count ::mlir::IntegerAttr 64-bit signless integer attribute
batch_group_count ::mlir::IntegerAttr 64-bit signless integer attribute
precision_config ::mlir::ArrayAttr Precision Config attribute

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.copy (::mlir::mhlo::CopyOp)

Copy operator

Syntax:

operation ::= `mhlo.copy` operands attr-dict `:` functional-type(operands, results)

Returns a copy of operand.

Example:

%0 = mhlo.copy %arg0 : (tensor<f32>) -> tensor<f32>

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.cosine (::mlir::mhlo::CosineOp)

Cos operator

Returns Cos(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.create_token (::mlir::mhlo::CreateTokenOp)

Create Token operator

Syntax:

operation ::= `mhlo.create_token` attr-dict `:` type(results)

Produces a HLO token. Tokens are used for ordering side-effecting operations. This is exported to HLO as an AfterAll operation with no operands to generate a token.

Example:

%1 = mhlo.create_token : !mhlo.token

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Results:

Result Description
output token

mhlo.cross-replica-sum (::mlir::mhlo::CrossReplicaSumOp)

Sums input across replicated instances.

For each of the replica groups, operands of the group devices are summed so that each device has the sum.

For example, suppose there are 8 TPU devices: [A, B, C, D, E, F, G, H]. Passing group_assignment=[[0,2,4,6],[1,3,5,7]] sets A, C, E, G as group 0, and B, D, F, H as group 1. Thus we get the outputs: [A+C+E+G, B+D+F+H, A+C+E+G, B+D+F+H, A+C+E+G, B+D+F+H, A+C+E+G, B+D+F+H].

See https://www.tensorflow.org/xla/operation_semantics#crossreplicasum

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
replica_groups ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.cstr_reshapable (::mlir::mhlo::CstrReshapableOp)

Compute input for reshape with any dynamic dim resolved

Syntax:

operation ::= `mhlo.cstr_reshapable` $num_elements `,` $dynamic_shape attr-dict `:` type($num_elements) `,` type($dynamic_shape)

This operation creates a witness on the constraint that a given shape would be a valid reshape for the given number of elements.

%0 = mhlo.cstr_reshapable 12, [2, -1] -> success
%1 = mhlo.cstr_reshapable 13, [2, -1] -> failure

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
num_elements index
dynamic_shape 1D tensor of integer or index values

Results:

Result Description
result

mhlo.custom_call (::mlir::mhlo::CustomCallOp)

CustomCall operator

A custom call invokes code external to XLA. The args are passed to the external code, and the external code is expected to produce a result of the given type. The exact mechanism is backend-specific. For example, in the CPU backend, a call instruction is emitted which targets a symbol with the name call_target_name.

call_target_name and backend_config can be arbitrary strings, but call_target_name should be short as it may be used in labels. backend_config can encode arbitrarily large amounts of information.

has_side_effect must be true if the custom call has side-effects. api_version specifies the version of the API used by the custom call function.

A custom call may apply functions within the scope of the parent module. They can be referenced using called_computations attribute.

A custom call can also have layout constraints on operands and results which can be specified as optional operand_layouts and result_layouts attributes. The layout attribute is an array of rank-1 index tensors and the i-th layout attribute specifies the layout for i-th operand/result.

The operand_layouts & result_layouts attributes can be specified under the following constraints:

1) Either both operand_layouts and result_layouts are specified or none. 2) None of the operands are of tuple type. 3) None of the results are of tuple type except the common case of single tuple result packing non-tuple values is allowed. In this case the i-th result_layouts attribute specifies the layout of i-th element in the result tuple.

See https://www.tensorflow.org/xla/operation_semantics#customcall

Interfaces: MemoryEffectOpInterface

Attributes:

Attribute MLIR Type Description
call_target_name ::mlir::StringAttr string attribute
has_side_effect ::mlir::BoolAttr bool attribute
backend_config ::mlir::StringAttr string attribute
api_version ::mlir::mhlo::CustomCallApiVersionAttr Custom call API version
called_computations ::mlir::ArrayAttr flat symbol ref array attribute
operand_layouts ::mlir::ArrayAttr Array of layout (1D tensor of index type) attributes
result_layouts ::mlir::ArrayAttr Array of layout (1D tensor of index type) attributes
output_operand_aliases ::mlir::ArrayAttr Aliasing attribute for outputs and operands of CustomCall

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.divide (::mlir::mhlo::DivOp)

Division operator

Returns lhs / rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.domain (::mlir::mhlo::DomainOp)

Marks groups of instructions (domains) with a property

Domain instructions are used to group instructions with the same DomainMetadata property. ShardingMetadata is the main use case today to group instructions on the same device. Domain instructions provide two major benefits:

  • Prevent unintentionally optimizing instructions across domains.
  • Automatically assign the metadata of the instructions created in the domain. Without domain instructions, each HLO optimization pass would have to check and propagate the metadata, which would be easy to miss and also adds complexity to the compiler. Since domain instructions connect two different domains, each domain instruction is associated with two DomainMetadata -- one on the operand side and one on the user side of the domain.

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
kind ::mlir::mhlo::DomainKindAttr Kind of domain metatdata attached to an HLO domain.
entry_metadata ::mlir::StringAttr string attribute
exit_metadata ::mlir::StringAttr string attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.dot_general (::mlir::mhlo::DotGeneralOp)

General Dot operator

Performs general dot products between vectors, vector/matrix and matrix/matrix multiplication.

See https://www.tensorflow.org/xla/operation_semantics#dotgeneral

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dot_dimension_numbers ::mlir::mhlo::DotDimensionNumbersAttr Attribute that models the dimension information for dot.
precision_config ::mlir::ArrayAttr Precision Config attribute

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dot (::mlir::mhlo::DotOp)

Dot operator

Performs dot products between vectors, vector/matrix and matrix/matrix multiplication.

See https://www.tensorflow.org/xla/operation_semantics#dot

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
precision_config ::mlir::ArrayAttr Precision Config attribute

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_broadcast_in_dim (::mlir::mhlo::DynamicBroadcastInDimOp)

Broadcast a tensor into the given dynamic shape by adding dimensions.

This is a generalization of the BroadcastInDimOp which accepts its output dimensions as an argument. It should eventually supercede the statically shaped original, but is being phased as a separate op in order to support compatibility with lowerings and translations that precede dynamic shapes.

The op accepts optional attributes to express static knowledge about the expanding behavior of dimensions. If not specified, all dimensions are assumed to be possibly expanding. The sets of dimensions that are known to be expanding and the set of dimensions that are known to be non-expanding must be disjoint and they must be a subset of the operand's dimensions.

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
broadcast_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
known_expanding_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
known_nonexpanding_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
output_dimensions 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_conv (::mlir::mhlo::DynamicConvOp)

Dynamic Convolution operator

The dynamic shape version of ConvOp. Computes a convolution with dynamic padding.

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
window_strides ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
padding ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
lhs_dilation ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
rhs_dilation ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
window_reversal ::mlir::DenseElementsAttr constant boolean vector/tensor attribute
dimension_numbers ::mlir::mhlo::ConvDimensionNumbersAttr Structure of dimension information for conv op
feature_group_count ::mlir::IntegerAttr 64-bit signless integer attribute
batch_group_count ::mlir::IntegerAttr 64-bit signless integer attribute
precision_config ::mlir::ArrayAttr Precision Config attribute

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
d_padding tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_gather (::mlir::mhlo::DynamicGatherOp)

Dynamic Gather operator

The dynamic shape version of GatherOp. Stitches together several slices of an input array.

Traits: InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimension_numbers ::mlir::mhlo::GatherDimensionNumbersAttr Attribute that models the dimension information for gather
indices_are_sorted ::mlir::BoolAttr bool attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
start_indices tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
slice_sizes tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_iota (::mlir::mhlo::DynamicIotaOp)

Create linear increasing values from 0 to length -1.

Produces an HLO Tensor of the specified shape, with an incremental set of values along the specified dimension starting at 0.

Requires:

  • The output length of the tensor result.

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
iota_dimension ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
output_shape 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_pad (::mlir::mhlo::DynamicPadOp)

Dynamic Pad operator

Syntax:

operation ::= `mhlo.dynamic_pad` operands attr-dict `:` functional-type(operands, results)

Dynamically Pads the operand, with amount of padding added at low-end/high-end/interior is passed through input tensors.

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
padding_value tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
edge_padding_low 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
edge_padding_high 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
interior_padding 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_reshape (::mlir::mhlo::DynamicReshapeOp)

Reshape a tensor to a given, possibly dynamic, shape.

Syntax:

operation ::= `mhlo.dynamic_reshape` operands attr-dict `:` functional-type(operands, results)

Reshapes operand to output_shape.

Requires:

  • The length of output_shape is equal to the rank of result.
  • The number of elements in operand (that is, the product of extents of its shape) is equal to the number of elements in output_shape (that is, the product of values in output_shape).

Example:

%0 = mhlo.dynamic_reshape %arg0, %shape : (tensor<?xf32>, tensor<2xindex>) -> tensor<?x?xf32>

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
output_shape 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_slice (::mlir::mhlo::DynamicSliceOp)

Dynamic Slice operator

Extracts a sub-array from the input array at dynamic start_indices.

See https://www.tensorflow.org/xla/operation_semantics#dynamicslice

Traits: InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
slice_sizes ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
start_indices 0D tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.dynamic_update_slice (::mlir::mhlo::DynamicUpdateSliceOp)

Dynamic Update Slice operator

Syntax:

operation ::= `mhlo.dynamic_update_slice` operands attr-dict `:` functional-type(operands, results)

DynamicUpdateSlice generates a result which is the value of the input array operand, with a slice update overwritten at start_indices.

See https://www.tensorflow.org/xla/operation_semantics#dynamicupdateslice

Example:

%0 = mhlo.dynamic_update_slice %arg0, %arg1, %arg2
       : (tensor<4xf32>, tensor<2xf32>, tensor<i32>) -> tensor<4xf32>

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
update tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
start_indices 0D tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.einsum (::mlir::mhlo::EinsumOp)

Einsum operator

Returns a tensor whose elements are defined by equation, which is written in a shorthand form inspired by the Einstein summation convention.

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
einsum_config ::mlir::StringAttr string attribute

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.exponential (::mlir::mhlo::ExpOp)

Exponential operator

Returns e^(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.exponential_minus_one (::mlir::mhlo::Expm1Op)

Exponential minus one operator

Returns e^(operand) - 1 element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.fft (::mlir::mhlo::FftOp)

Fast fourier transform operator

Returns the fast-fourier-transform of the input array.

See https://www.tensorflow.org/xla/operation_semantics#fft

Traits: InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
fft_type ::mlir::mhlo::FftTypeAttr XLA fast fourier transform type.
fft_length ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.floor (::mlir::mhlo::FloorOp)

Floor operator

Returns Floor(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.fusion (::mlir::mhlo::FusionOp)

Fusion operator

Models the fusion instruction.

A fusion op is consists of a group of basic ops (represented as a region attached to it). It serves as a hint to the backend that it is beneficial to emit the contained ops into a single loop nest or kernel.

Attributes:

Attribute MLIR Type Description
fusion_kind ::mlir::mhlo::FusionKindAttr fusion kind

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

Results:

Result Description
results tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.gather (::mlir::mhlo::GatherOp)

Gather operator

Stitches together several slices of operand from offsets specified in start_indices (each slice at a potentially different runtime offset).

See https://www.tensorflow.org/xla/operation_semantics#gather

Traits: InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimension_numbers ::mlir::mhlo::GatherDimensionNumbersAttr Attribute that models the dimension information for gather
slice_sizes ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
indices_are_sorted ::mlir::BoolAttr bool attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
start_indices tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.get_dimension_size (::mlir::mhlo::GetDimensionSizeOp)

GetDimensionSize operator

Returns the size of the given dimension of the operand.

See https://www.tensorflow.org/xla/operation_semantics#getdimensionsize

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimension ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 32-bit signless integer values

mhlo.get_tuple_element (::mlir::mhlo::GetTupleElementOp)

GetTupleElement operator

Returns a member of a tuple specified by an index.

See https://www.tensorflow.org/xla/operation_semantics#gettupleelement

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
index ::mlir::IntegerAttr 32-bit signless integer attribute

Operands:

Operand Description
«unnamed» nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.if (::mlir::mhlo::IfOp)

If operator

Executes the function true_branch if pred is true or false_branch if pred is false, and returns the result.

The type of the returned values of true_branch and false_branch functions must be the same and equal to the types of the values returned by the operation.

Note that only one of two functions will be executed depending on the value of pred.

Traits: RecursiveSideEffects, SingleBlockImplicitTerminator

Operands:

Operand Description
pred tensor of pred (AKA boolean or 1-bit integer) values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.imag (::mlir::mhlo::ImagOp)

Imag operator

Returns Imag(operand) element-wise.

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.infeed (::mlir::mhlo::InfeedOp)

Infeed operator

Reads a single data item from the implicit Infeed streaming interface of the device, interpreting the data as the given shape, and returns a XlaOp of the data. Multiple Infeed operations are allowed in a computation, but there must be a total order among the Infeed operations.

Attributes: layout: Array attribute. Each element of the array is a minor_to_major array corresponding to the shape of the data read from the infeed interface.

See https://www.tensorflow.org/xla/operation_semantics#infeed

Attributes:

Attribute MLIR Type Description
infeed_config ::mlir::StringAttr string attribute
layout ::mlir::ArrayAttr array attribute

Operands:

Operand Description
token token

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.iota (::mlir::mhlo::IotaOp)

Iota operator

Creates a rank 1 array of values starting at zero and incrementing by one.

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
iota_dimension ::mlir::IntegerAttr 64-bit signless integer attribute

Results:

Result Description
output tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.is_finite (::mlir::mhlo::IsFiniteOp)

IsFinite operator

Tests whether each element of operand is finite, i.e., is not positive or negative infinity, and is not NaN. Returns a tensor of 1-bit integers with the same shape as the input, where each element is nonzero (i.e. true) if and only if the corresponding input element is finite.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
x tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
y tensor of pred (AKA boolean or 1-bit integer) values

mhlo.log_plus_one (::mlir::mhlo::Log1pOp)

Log1p operator

Returns log(operand+1) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.log (::mlir::mhlo::LogOp)

Logarithm operator

Returns log(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.logistic (::mlir::mhlo::LogisticOp)

Logistic operator

Returns logistic(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.map (::mlir::mhlo::MapOp)

Map operator

Applies a scalar function over the given operands arrays, producing an array of the same dimensions where each element is the result of the mapped function applied to the corresponding elements in the input arrays.

The mapped function is an arbitrary computation with the restriction that it has N inputs of scalar type T and a single output with type S. The output has the same dimensions as the operands except that the element type T is replaced with S.

See https://www.tensorflow.org/xla/operation_semantics#map

Traits: RecursiveSideEffects, SameOperandsAndResultShape, SingleBlockImplicitTerminator

Interfaces: InferShapedTypeOpInterface

Attributes:

Attribute MLIR Type Description
dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.maximum (::mlir::mhlo::MaxOp)

Maximum operator

Returns max(lhs, rhs) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.minimum (::mlir::mhlo::MinOp)

Minimum operator

Returns min(lhs, rhs) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.multiply (::mlir::mhlo::MulOp)

Multiplication operator

Returns lhs * rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.negate (::mlir::mhlo::NegOp)

Negation operator

Returns -operand element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.not (::mlir::mhlo::NotOp)

Not operator

Returns biwise-NOT of operand element-wise. The input tensor must be of type integer HLO_Int or boolean HLO_Pred.

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

mhlo.optimization_barrier (::mlir::mhlo::OptimizationBarrierOp)

The `mhlo.optimization_barrier` op blocks optimizations.

Example:

```mlir
%0:2 = mhlo.optimization_barrier %arg0, %arg1 : (tensor<4x4xf32>, tensor<3x4xf32>) -> (tensor<4x4xf32>, tensor<3x4xf32>)
```

Syntax:

operation ::= `mhlo.optimization_barrier` operands attr-dict `:` functional-type(operands, results)

Blocks any optimization pass from moving computations across the barrier.

Ensures that all inputs are evaluated before any operators that depend on the barrier's outputs. See https://www.tensorflow.org/xla/operation_semantics#optimizationbarrier

Traits: HLO_PairwiseSameOperandAndResultType

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.or (::mlir::mhlo::OrOp)

Or operator

Returns biwise-OR of lhs and rhs element-wise. The input tensors must be of type integer HLO_Int or boolean HLO_Pred.

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
rhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.outfeed (::mlir::mhlo::OutfeedOp)

Outfeed operator

Generates outgoing data transfers for the given data. It takes data and a token type operand and produces a token type value. Tokens are used for ordering side-effecting operations.

See https://www.tensorflow.org/xla/operation_semantics#outfeed

Attributes:

Attribute MLIR Type Description
outfeed_config ::mlir::StringAttr string attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
token token

Results:

Result Description
«unnamed» token

mhlo.pad (::mlir::mhlo::PadOp)

Pad operator

Pads edges and between the elements of operand with the padding_value according to the configuration parameters described below.

edge_padding_low and edge_padding_high specify the amount of padding added at the low-end (next to index 0) and the high-end (next to the highest index) of each dimension respectively. The amount of edge padding can be negative -- the absolute value of negative padding indicates the number of elements to remove from the specified dimension.

interior_padding specifies the amount of padding (non-negative) added between any two elements in each dimension. Interior padding occurs logically before edge padding, so in the case of negative edge padding, elements are removed from the interior-padded operand.

This operation is a no-op if, for all dimensions, the edge padding pairs are all (0, 0) and the interior padding values are all 0. The figure below shows examples of different edge_padding and interior_padding values for a two-dimensional array.

Examples

Traits: InferTensorType, SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
edge_padding_low ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
edge_padding_high ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
interior_padding ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
padding_value tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.partition_id (::mlir::mhlo::PartitionIdOp)

PartitionId operator

Syntax:

operation ::= `mhlo.partition_id` attr-dict `:` type(results)

Returns the value of the partition id of the currently executing device. XLA supports two mechanisms for parallel execution: replication and partition. A module can be replicated to run on multiple devices (replicas) and a module can also be partitioned to split the work between devices. replica-id and partition-id returns the id values of the current device.

Example:

%1 = mhlo.partition_id : tensor<ui32>

Results:

Result Description
«unnamed» tensor of 32-bit unsigned integer values

mhlo.popcnt (::mlir::mhlo::PopulationCountOp)

PopulationCount operator

Returns the number of bits set in each operand element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

mhlo.power (::mlir::mhlo::PowOp)

Power operator

Returns lhs ^ rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.real_dynamic_slice (::mlir::mhlo::RealDynamicSliceOp)

Real Dynamic Slice operator

Syntax:

operation ::= `mhlo.real_dynamic_slice` operands attr-dict `:` functional-type(operands, results)

The dynamic shape version of SliceOp. Extracts a sub-array from the input array according to start_indices, limit_indices and strides. Expect start_indices/limit_indices/strides to be statically shaped and matching the rank of the input.

Example:

%0 = mhlo.real_dynamic_slice %input, %start, %limit, %strides
       : (tensor<256x?xf32>, tensor<2xindex>, tensor<2xindex>, tensor<2xindex>) -> tensor<256x?xf32>

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
start_indices 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
limit_indices 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
strides 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.real (::mlir::mhlo::RealOp)

Real operator

Returns Real(operand) element-wise.

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.recv (::mlir::mhlo::RecvOp)

Recv operator

Receives data of the given shape from a Send instruction in another computation that shares the same channel handle. Returns a tuple containing value for the received data and a token. Recv operation represents synchronous communication. However, the instruction is internally decomposed into 2 HLO instructions (Recv and RecvDone) to enable asynchronous data transfers.

See https://www.tensorflow.org/xla/operation_semantics#recv

Attributes:

Attribute MLIR Type Description
channel_handle ::mlir::mhlo::ChannelHandleAttr two 64-bit integers 'handle' and 'type'
is_host_transfer ::mlir::BoolAttr bool attribute

Operands:

Operand Description
token token

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.reduce (::mlir::mhlo::ReduceOp)

Reduce operator

Returns the result of executing a reduction function on one or more arrays in parallel.

See https://www.tensorflow.org/xla/operation_semantics#reduce

Traits: RecursiveSideEffects, SameVariadicOperandSize, SingleBlockImplicitTerminator

Interfaces: InferShapedTypeOpInterface

Attributes:

Attribute MLIR Type Description
dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
init_values tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.reduce_precision (::mlir::mhlo::ReducePrecisionOp)

Reduce precision operator

Models the effect of converting floating - point values to a lower - precision format(such as IEEE - FP16) and back to the original format. The number of exponent and mantissa bits in the lower - precision format can be specified arbitrarily, although all bit sizes may not be supported on all hardware implementations.

See https://www.tensorflow.org/xla/operation_semantics#reduceprecision

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
exponent_bits ::mlir::IntegerAttr 32-bit signless integer attribute
mantissa_bits ::mlir::IntegerAttr 32-bit signless integer attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
output tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.reduce_scatter (::mlir::mhlo::ReduceScatterOp)

ReduceScatter operator

Performs all_reduce followed by a scatter.

See https://www.tensorflow.org/xla/operation_semantics#reducescatter

Traits: SameOperandsAndResultElementType

Attributes:

Attribute MLIR Type Description
scatter_dimension ::mlir::IntegerAttr 64-bit signless integer attribute
replica_groups ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
channel_handle ::mlir::mhlo::ChannelHandleAttr two 64-bit integers 'handle' and 'type'
use_global_device_ids ::mlir::UnitAttr unit attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.reduce_window (::mlir::mhlo::ReduceWindowOp)

ReduceWindow operator

Returns the result of executing a reduction function over all elements in each window of one or more arrays in parallel.

See https://www.tensorflow.org/xla/operation_semantics#reducewindow

Traits: RecursiveSideEffects, SameVariadicOperandSize, SingleBlockImplicitTerminator

Attributes:

Attribute MLIR Type Description
window_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
window_strides ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
base_dilations ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
window_dilations ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
padding ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
init_values tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.remainder (::mlir::mhlo::RemOp)

Remainder operator

Returns lhs % rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.replica_id (::mlir::mhlo::ReplicaIdOp)

ReplicaId operator

Syntax:

operation ::= `mhlo.replica_id` attr-dict `:` type(results)

Returns the unique ID (int32 scalar) of the replica.

The unique ID of each replica is an unsigned integer in the interval [0, N), where N is the number of replicas. Since all the replicas are running the same program, a ReplicaId() call in the program will return a different value on each replica.

See https://www.tensorflow.org/xla/operation_semantics#replicaid

Example:

%0 = mhlo.replica_id : tensor<ui32>

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Results:

Result Description
«unnamed» tensor of 32-bit unsigned integer values

mhlo.reshape (::mlir::mhlo::ReshapeOp)

Reshape operator

Syntax:

operation ::= `mhlo.reshape` operands attr-dict `:` functional-type(operands, results)

Reshapes the dimensions of operand into a new configuration.

See https://www.tensorflow.org/xla/operation_semantics#reshape

Example:

%0 = mhlo.reshape %arg0 : (tensor<2xf32>) -> tensor<1x2xf32>

Traits: SameOperandsAndResultElementType

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» statically shaped tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.return (::mlir::mhlo::ReturnOp)

The `hlo.return` operation terminates a region and returns values.

Example:

```mlir
%0 = mhlo.reduce %arg0, %arg1 {
  ...
  mhlo.return %1 : tensor<f32>
}
```

Syntax:

operation ::= `mhlo.return` $results attr-dict (`:` type($results)^)?

Traits: Terminator

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
results tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.reverse (::mlir::mhlo::ReverseOp)

Reverse operator

Reverses the specified dimensions of operand according to the given dimensions.

See https://www.tensorflow.org/xla/operation_semantics#rev_reverse

Traits: CompatibleOperandsAndResultType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.rng_bit_generator (::mlir::mhlo::RngBitGeneratorOp)

Uniform random number generator operator

Returns an output with a given shape filled with uniform random bits using the specified algorithm (or backend default) and returns an updated state (with the same shape as initial state) and the generated random data.

See https://www.tensorflow.org/xla/operation_semantics#rngbitgenerator

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
rng_algorithm ::mlir::mhlo::RngAlgorithmAttr XLA PRNG algorithm to be used.

Operands:

Operand Description
initial_state tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
output_state tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
output tensor of 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.rng (::mlir::mhlo::RngOp)

RNG with uniform distribution.

Constructs an output of a given shape with random numbers generated following the given rng_distribution with two parameters: UNIFORM: the uniform distribution over the interval [a,b). The parameters and output element type have to be a boolean type, an integral type or a floating point types, and the types have to be consistent.

         See <a href="https://www.tensorflow.org/xla/operation_semantics#rnguniform">https://www.tensorflow.org/xla/operation_semantics#rnguniform</a>

NORMAL: the normal distribution with parameters mu (=a) and sigma (=b). The parameters and output shape have to have a floating point elemental type. The parameters furthermore have to be scalar valued.

        See <a href="https://www.tensorflow.org/xla/operation_semantics#rngnormal">https://www.tensorflow.org/xla/operation_semantics#rngnormal</a>

Traits: InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
rng_distribution ::mlir::mhlo::RngDistributionAttr XLA PRNG distribution to be used.

Operands:

Operand Description
a 0D tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
b 0D tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values
shape 1D tensor of index or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.round_nearest_even (::mlir::mhlo::RoundNearestEvenOp)

Round operator, ties to even

Returns Round(operand) element-wise, rounding to nearest integer with half-way cases rounding towards even numbers.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.round_nearest_afz (::mlir::mhlo::RoundOp)

Round operator, ties away from zero

Returns Round(operand) element-wise, rounding to nearest integer with half-way cases rounding away from zero.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type values

mhlo.rsqrt (::mlir::mhlo::RsqrtOp)

Reciprocal Square-root operator

Returns 1.0 / sqrt(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.scatter (::mlir::mhlo::ScatterOp)

Scatter operator

Generates a result which is the value of the input array operand, with several slices (at indices specified by scatter_indices) updated with the values in updates using update_computation.

See https://www.tensorflow.org/xla/operation_semantics#scatter

Traits: RecursiveSideEffects, SameVariadicOperandSize

Attributes:

Attribute MLIR Type Description
scatter_dimension_numbers ::mlir::mhlo::ScatterDimensionNumbersAttr Attribute that models the dimension information for scatter
indices_are_sorted ::mlir::BoolAttr bool attribute
unique_indices ::mlir::BoolAttr bool attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
scatter_indices tensor of integer or index values
updates tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.select_and_scatter (::mlir::mhlo::SelectAndScatterOp)

SelectAndScatter operator

Runs a windowed selection select function over operand with shape window_dimensions and stride window_strides. This will produce an amount of selected locations whose shape matches source. These are then scattered to the output which is initialized with init_value. Multiple scattered elements which land in the same output location are combined using the scatter function.

See https://www.tensorflow.org/xla/operation_semantics#selectandscatter

Traits: RecursiveSideEffects

Attributes:

Attribute MLIR Type Description
window_dimensions ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
window_strides ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
padding ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
source tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
init_value tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.select (::mlir::mhlo::SelectOp)

Select operator

Constructs an output tensor from the elements of on_true and on_false based on the values of pred. All three operands must be of the same shape with the exception of pred, which may also be a scalar in which case it is broadcasted.

See https://www.tensorflow.org/xla/operation_semantics#select

Traits: HLO_BroadcastingElementwise, InferTensorType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
pred tensor of pred (AKA boolean or 1-bit integer) values
on_true tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
on_false tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.send (::mlir::mhlo::SendOp)

Send operator

Sends the given operand data to a Recv instruction in another computation that shares the same channel handle. Does not return any data. Similar to the Recv operation, Send operation represents synchronous communication, and is internally decomposed into 2 HLO instructions (Send and SendDone) to enable asynchronous data transfers.

See https://www.tensorflow.org/xla/operation_semantics#send

Attributes:

Attribute MLIR Type Description
channel_handle ::mlir::mhlo::ChannelHandleAttr two 64-bit integers 'handle' and 'type'
is_host_transfer ::mlir::BoolAttr bool attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
token token

Results:

Result Description
«unnamed» token

mhlo.set_dimension_size (::mlir::mhlo::SetDimensionSizeOp)

SetDimensionSize operator

Sets the dynamic size of operand's given dimension. Pass through the operand as result, with dynamic dimension tracked by the compiler. Padded values will be ignored by downstream reduction ops.

See https://www.tensorflow.org/xla/operation_semantics#setdimensionsize

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dimension ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
size tensor of 32-bit signless integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.shift_left (::mlir::mhlo::ShiftLeftOp)

Shift Left operator

Returns lhs << rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.shift_right_arithmetic (::mlir::mhlo::ShiftRightArithmeticOp)

Shift right arithmetic operator

Returns arithmetic lhs >> rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.shift_right_logical (::mlir::mhlo::ShiftRightLogicalOp)

Shift right logical operator

Returns logical lhs >> rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.sign (::mlir::mhlo::SignOp)

Sign operator

Returns sign(operand) element-wise, where

sign(x) = -1  : x < 0
        = -0  : x = -0
        = NaN : x = NaN
        = +0  : x = +0
        = 1   : x > 0

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32/64-bit signless integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 4/8/16/32/64-bit signless integer or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.sine (::mlir::mhlo::SineOp)

Sin operator

Returns Sin(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.slice (::mlir::mhlo::SliceOp)

Traits: SameOperandsAndResultElementType

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
start_indices ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
limit_indices ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute
strides ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.sort (::mlir::mhlo::SortOp)

Sort operator

Sorts the given operands at the given dimension with the given comparator.

See https://www.tensorflow.org/xla/operation_semantics#sort

Traits: RecursiveSideEffects, SameOperandsAndResultShape

Attributes:

Attribute MLIR Type Description
dimension ::mlir::IntegerAttr 64-bit signless integer attribute
is_stable ::mlir::BoolAttr bool attribute

Operands:

Operand Description
operands tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.sqrt (::mlir::mhlo::SqrtOp)

Square-root operator

Returns sqrt(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.subtract (::mlir::mhlo::SubtractOp)

Subtraction operator

Returns lhs - rhs element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_binary_arithmetic_operations

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
rhs tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.tanh (::mlir::mhlo::TanhOp)

Tanh operator

Returns tanh(operand) element-wise.

See https://www.tensorflow.org/xla/operation_semantics#element-wise_unary_functions

Traits: CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.torch_index_select (::mlir::mhlo::TorchIndexSelectOp)

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
dim ::mlir::IntegerAttr 64-bit signless integer attribute
batch_dims ::mlir::IntegerAttr 64-bit signless integer attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values
index tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.trace (::mlir::mhlo::TraceOp)

Trace operator

Syntax:

operation ::= `mhlo.trace` $operand `,` $tag attr-dict `:` type($operand)

Emits a logging message tag with the operand.

Example:

mhlo.trace %arg0, "In test code." : tensor<5x1x5xi32>

Attributes:

Attribute MLIR Type Description
tag ::mlir::StringAttr string attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.transpose (::mlir::mhlo::TransposeOp)

Transpose operator

Permutes the dimensions of operand according to the given permutation.

res_dimensions[i] = operand_dimensions[permutation[i]]

See https://www.tensorflow.org/xla/operation_semantics#transpose

Traits: SameOperandsAndResultElementType

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
permutation ::mlir::DenseIntElementsAttr 64-bit signless integer elements attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.triangular_solve (::mlir::mhlo::TriangularSolveOp)

TriangularSolve operator

Solves systems of linear equations with lower or upper triangular coefficient matrices by forward- or back-substitution. Broadcasting along leading dimensions, this routine solves one of the matrix systems op(a) * x = b, or x * op(a) = b, for the variable x, given a and b, where op(a) is either op(a) = a, or op(a) = Transpose(a), or op(a) = Conj(Transpose(a)).

Input data is read only from the lower/upper triangle of a, depending on the value of lower. Values from the other triangle are ignored. Output data is returned in the same triangle; the values in the other triangle are implementation-defined and may be anything.

If the rank of a and b are greater than 2, they are treated as batches of matrices, where all except the minor 2 dimensions are batch dimensions. a and b must have equal batch dimensions.

See https://www.tensorflow.org/xla/operation_semantics#triangularsolve

Traits: SameOperandsAndResultElementType

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
left_side ::mlir::BoolAttr bool attribute
lower ::mlir::BoolAttr bool attribute
unit_diagonal ::mlir::BoolAttr bool attribute
transpose_a ::mlir::mhlo::TransposeAttr Transpose options

Operands:

Operand Description
a tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values
b tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or complex type with 32-bit float or 64-bit float elements values

mhlo.tuple (::mlir::mhlo::TupleOp)

XLA's tuple op

Groups a set of tensor inputs into a single tuple object.

See https://www.tensorflow.org/xla/operation_semantics#tuple

Interfaces: InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
val tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token or nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

Results:

Result Description
«unnamed» nested tuple with any combination of tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token values

mhlo.unary_einsum (::mlir::mhlo::UnaryEinsumOp)

Einsum operator

Returns a tensor whose elements are defined by equation, which is written in a shorthand form inspired by the Einstein summation convention.

Interfaces: NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Attributes:

Attribute MLIR Type Description
einsum_config ::mlir::StringAttr string attribute

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.uniform_dequantize (::mlir::mhlo::UniformDequantizeOp)

Uniform dequantize operator

Converts quantized array of integers to floating-points according to the quantization parameters defined by the input type.

Example:

%0 = mhlo.uniform_dequantize %arg0 : (tensor<16x16x!quant.uniform<i8:f32, 34.0:16>>) -> tensor<16x16xf32>

Traits: Elementwise, InferTensorType, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 32-bit float or bfloat16 type values

mhlo.uniform_quantize (::mlir::mhlo::UniformQuantizeOp)

Uniform quantize operator

Converts floating point tensors or uniform quantized integer tensors to uniform quantized integer tensors according to the quantization parameters defined by the output type.

Example:

%0 = mhlo.uniform_quantize %arg0 : (tensor<16x16xf32>) -> tensor<16x16x!quant.uniform<ui8:f32, 34.0:16>>

Traits: Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
operand tensor of 32-bit float or bfloat16 type or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Results:

Result Description
result tensor of 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

mhlo.while (::mlir::mhlo::WhileOp)

While operator

Returns the result of executing a body function until the cond body returns true.

See https://www.tensorflow.org/xla/operation_semantics#while

Traits: HLO_PairwiseSameOperandAndResultType, RecursiveSideEffects, SingleBlockImplicitTerminator

Interfaces: OpAsmOpInterface

Operands:

Operand Description
operand tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

Results:

Result Description
«unnamed» tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values or token

mhlo.xla.rng_get_and_update_state (::mlir::mhlo::XlaRngGetAndUpdateStateOp)

RNG state change

Syntax:

operation ::= `mhlo.xla.rng_get_and_update_state` attr-dict

This instruction represents the change of the global random number generator state for rng instructions. The global state is incremented by delta and the old state is returned.

The output is currently defined for a single output type. If this changes in the future to support multiple types, lowering to use of a global memref must ensure that a single memref is still used and updated appropriately.

Interfaces: InferTypeOpInterface

Attributes:

Attribute MLIR Type Description
delta ::mlir::IntegerAttr 64-bit signless integer attribute

Results:

Result Description
«unnamed» statically shaped tensor of 64-bit unsigned integer values

mhlo.xor (::mlir::mhlo::XorOp)

Xor operator

Returns biwise-XOR of lhs and rhs element-wise. The input tensors must be of type integer HLO_Int or boolean HLO_Pred.

Traits: Commutative, CompatibleOperandsAndResultType, Elementwise, SameOperandsAndResultShape

Interfaces: InferShapedTypeOpInterface, InferTypeOpInterface, NoSideEffect (MemoryEffectOpInterface)

Effects: MemoryEffects::Effect{}

Operands:

Operand Description
lhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values
rhs tensor of pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer values

Results:

Result Description
result tensor of 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 4/8/16/32/64-bit signless integer or 4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 4/8/16/32-bit uniform quantized signed integer or 4/8/16/32-bit uniform quantized unsigned integer values

Attribute definition

ArgResultAliasAttr

Attribute that models the alias relationship of entry function argument

This attribute captures the alias relationship of an MHLO main function argument to one of the results, denoted by resultIndex. The argTupleIndices and resultTupleIndices are used to index into nested tuples in operand and result respectively. If isMustAlias is true then the operand-result pair must alias.

This is meant to be used as an attribute on a function argument in MHLO. For example, in the following code it expresses that %arg1 may alias 0-th result.

func @main(%arg0: tensor<2xf32>, %arg1: tensor<3xf32> {mhlo.result_alias =
    mhlo.result_alias<result_index = [2], ...>}
  ) -> tensor<2xf32>, tensor<3xf32> {
  // function body ...
}

Parameters:

Parameter C++ type Description
argTupleIndices ::llvm::ArrayRef<int64_t> Dimension
resultIndex int64_t
resultTupleIndices ::llvm::ArrayRef<int64_t> Dimension
isMustAlias bool

ChannelHandleAttr

two 64-bit integers 'handle' and 'type'

Syntax:

!mhlo.channel_handle<
  int64_t,   # handle
  int64_t   # type
>

Parameters:

Parameter C++ type Description
handle int64_t
type int64_t

ConvDimensionNumbersAttr

Structure of dimension information for conv op

Parameters:

Parameter C++ type Description
inputBatchDimension int64_t
inputFeatureDimension int64_t
inputSpatialDimensions ::llvm::ArrayRef<int64_t> Dimension
kernelInputFeatureDimension int64_t
kernelOutputFeatureDimension int64_t
kernelSpatialDimensions ::llvm::ArrayRef<int64_t> Dimension
outputBatchDimension int64_t
outputFeatureDimension int64_t
outputSpatialDimensions ::llvm::ArrayRef<int64_t> Dimension

DotDimensionNumbersAttr

Attribute that models the dimension information for dot.

Parameters:

Parameter C++ type Description
lhsBatchingDimensions ::llvm::ArrayRef<int64_t> Dimension
rhsBatchingDimensions ::llvm::ArrayRef<int64_t> Dimension
lhsContractingDimensions ::llvm::ArrayRef<int64_t> Dimension
rhsContractingDimensions ::llvm::ArrayRef<int64_t> Dimension

GatherDimensionNumbersAttr

Attribute that models the dimension information for gather

Parameters:

Parameter C++ type Description
offsetDims ::llvm::ArrayRef<int64_t> Dimension
collapsedSliceDims ::llvm::ArrayRef<int64_t> Dimension
startIndexMap ::llvm::ArrayRef<int64_t> Dimension
indexVectorDim int64_t

ComparisonDirectionAttr

Which comparison operation to perform.

Syntax:

!mhlo.comparison_direction<
  ::mlir::mhlo::ComparisonDirection   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::ComparisonDirection an enum of type ComparisonDirection

ComparisonTypeAttr

Which comparison type to use.

Syntax:

!mhlo.comparison_type<
  ::mlir::mhlo::ComparisonType   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::ComparisonType an enum of type ComparisonType

DequantizeModeAttr

Dequantization mode. Only MIN_COMBINED is supported.

Syntax:

!mhlo.dequantize_mode<
  ::mlir::mhlo::DequantizeMode   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::DequantizeMode an enum of type DequantizeMode

DomainKindAttr

Kind of domain metatdata attached to an HLO domain.

Syntax:

!mhlo.kind<
  ::mlir::mhlo::DomainKind   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::DomainKind an enum of type DomainKind

FftTypeAttr

XLA fast fourier transform type.

Syntax:

!mhlo.fft_type<
  ::mlir::mhlo::FftType   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::FftType an enum of type FftType

FusionKindAttr

fusion kind

Syntax:

!mhlo.fusion_kind<
  ::mlir::mhlo::FusionKind   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::FusionKind an enum of type FusionKind

PrecisionAttr

XLA precision for an operand. Has backend specific meaning.

Syntax:

!mhlo.precision<
  ::mlir::mhlo::Precision   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::Precision an enum of type Precision

RngAlgorithmAttr

XLA PRNG algorithm to be used.

Syntax:

!mhlo.rng_algorithm<
  ::mlir::mhlo::RngAlgorithm   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::RngAlgorithm an enum of type RngAlgorithm

RngDistributionAttr

XLA PRNG distribution to be used.

Syntax:

!mhlo.rng_distribution<
  ::mlir::mhlo::RngDistribution   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::RngDistribution an enum of type RngDistribution

TransposeAttr

Transpose options

Syntax:

!mhlo.transpose<
  ::mlir::mhlo::Transpose   # value
>

Parameters:

Parameter C++ type Description
value ::mlir::mhlo::Transpose an enum of type Transpose

OutputOperandAliasAttr

Attribute that models the alias relationship of output and operand of a CustomCall op

Syntax:

!mhlo.output_operand_alias<
  ::llvm::ArrayRef<int64_t>,   # outputTupleIndices
  int64_t,   # operandIndex
  ::llvm::ArrayRef<int64_t>   # operandTupleIndices
>

This attribute captures the alias relationship of the output to one of the operands for a CustomCall op, denoted by operand_index. The output_tuple_indices and operand_tuple_indices are used to index into output and operand types. These indices lists are empty if the corresponding types are not tuple types, and can be arbitrarily long in case of arbitrarily nested tuple types.

See https://www.tensorflow.org/xla/aliasing

Example when used as array with in mhlo.custom-call:

%0 = "mhlo.custom_call"(%arg0, %arg1) {
  // other attributes
  output_operand_alias = [
    #mhlo.output_operand_alias<output_tuple_indices = [0],
                               operand_index = 0,
                               operand_tuple_indices = [1]>
  ]
} : (tuple<tensor<1x1xf32>, tensor<2x3xf32>>, tensor<5x5xf32>) -> tuple<tensor<2x3xf32>>

The output and the 0th operand are both tuples. The aliasing shows the
relationship between the 0th element in output tuple with the 1st element in
the 0th operand. And both of them are of the same type: tensor<2x3xf32>.

Parameters:

Parameter C++ type Description
outputTupleIndices ::llvm::ArrayRef<int64_t> Dimension
operandIndex int64_t
operandTupleIndices ::llvm::ArrayRef<int64_t> Dimension

ScatterDimensionNumbersAttr

Attribute that models the dimension information for scatter

Parameters:

Parameter C++ type Description
updateWindowDims ::llvm::ArrayRef<int64_t> Dimension
insertedWindowDims ::llvm::ArrayRef<int64_t> Dimension
scatterDimsToOperandDims ::llvm::ArrayRef<int64_t> Dimension
indexVectorDim int64_t

TypeExtensionsAttr

Attribute that extends tensor type with MHLO type properties.

Syntax:

!mhlo.type_extensions<
  ::llvm::ArrayRef<int64_t>   # bounds
>

This attribute is used to extend MLIR tensor type with MHLO tensor specific properties. These properties aren't modeled in the MLIR type. This attribute is set in the encoding field of the tensor type.

See HLO_BoundedAttrInterface for documentation for bounds.

Parameters:

Parameter C++ type Description
bounds ::llvm::ArrayRef<int64_t>

Type definition

AsyncBundleType

Opaque collection of other types

Syntax:

!mhlo.async_bundle<
  ::llvm::ArrayRef<Type>   # types
>

Parameters:

Parameter C++ type Description
types ::llvm::ArrayRef<Type>