פעולות
mhlo.abs
(mhlo::AbsOp)
ניתוח שרירי הבטן
תַחבִּיר:
operation ::= `mhlo.abs` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
מבצעת פעולת abs אלמנט על טנסור operand
ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#abs
דוּגמָה:
%result = mhlo.abs %operand : tensor<3xi32>
תכונות: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנזור מדורג של 2/4/8/16/32/64 סיביות חסר סימנים או סוג f4E2M1FN או סוג f6E2M3FN או סוג f6E3M2FN או סוג f8E3M4 או סוג f8E4M3 או סוג f8E4M3FN או סוג f8E4M3FNUZ או f8E4M3FNUZ type או f28EM3 או f28EM3 סוג FNUZ או f8E8M0FNU סוג או 16-bit float או 32-bit float או 64-bit float או bfloat16 סוג או סוג מורכב עם 32-bit float או 64-bit float אלמנטים או 2/4/8/16/32-bit סימן שלם אחיד או 2/4/8/16/32 סיביות אחיד מכומתי לכל ציר סימן שלם או 2/4/8/16/32 סיביות אחיד מכומד ללא סימן או 2/4/8/16/32 סיביות אחיד מכומד לכל ציר ללא סימן ערכים שלמים |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנזור מדורג של 2/4/8/16/32/64 סיביות חסר סימנים או סוג f4E2M1FN או סוג f6E2M3FN או סוג f6E3M2FN או סוג f8E3M4 או סוג f8E4M3 או סוג f8E4M3FN או סוג f8E4M3FNUZ או f8E4M3FNUZ type או f28EM3 או f28EM3 סוג FNUZ או f8E8M0FNU סוג או 16-bit float או 32-bit float או 64-bit float או bfloat16-סוג או 2/4/8/16/32-bit אחיד מרוכז בסימן שלם או 2/4/8/16/32-bit אחיד בכימות לפי מספר שלם בסימן ציר או 2/4/8/16/32 סיביות אחיד מכומד ללא סימן או מספר שלם שלם ב-2/4/8/16/32 סיביות אחיד בכימות לכל ציר ערכים ללא סימן |
mhlo.add
(mhlo::AddOp)
הוסף פעולה
תַחבִּיר:
operation ::= `mhlo.add` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
מבצע הוספת אלמנט של שני טנסורים lhs
ו- rhs
ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#add
דוּגמָה:
%result = mhlo.add %lhs, %rhs : tensor<2x2xi32>
תכונות: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
lhs | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16 /32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכים שלמים ללא סימן |
rhs | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16 /32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכים שלמים ללא סימן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16 /32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכים שלמים ללא סימן |
mhlo.add_dependency
(mhlo::AddDependencyOp)
פעולת AddDependency
תַחבִּיר:
operation ::= `mhlo.add_dependency` operands attr-dict `:` functional-type(operands, results)
פעולה זו פרטית ל-XLA מהדר, ולכן אין לה עדיין מפרט.
באופן לא רשמי, הפעולה הזו שני אופרנדים: אופרנד נתונים ואסימון. הפלט של הפעולה הוא אופרנד הנתונים. בשימוש עם AfterAll פעולה זו מאפשרת הזמנת פעולות שאינן משפיעות לוואי (אלו שאינן מייצרות ערכי אסימון).
דוּגמָה:
%1 = mhlo.add_dependency %arg0, %0 : (tensor<3x4xf32>, !mhlo.token) -> tensor<3x4xf32>
תכונות: AlwaysSpeculatableImplTrait
ממשקים: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי ציפה של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה ערכי מספר שלם ללא סימן או טנסור מדורג של 2/4 /8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן ציר או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכים או אסימון שלם ללא סימן |
token | אֲסִימוֹן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי ציפה של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה ערכי מספר שלם ללא סימן או טנסור מדורג של 2/4 /8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן ציר או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכים או אסימון שלם ללא סימן |
mhlo.after_all
(mhlo::AfterAllOp)
מבצע אחרי הכל
תַחבִּיר:
operation ::= `mhlo.after_all` $inputs attr-dict
`:` custom<VariadicSameOperandsAndResultType>(ref($inputs), type($inputs), type($result))
מבטיח שהפעולות המייצרות את inputs
מבוצעות לפני כל פעולות התלויות result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#after_all
דוּגמָה:
%result = mhlo.after_all %input0, %input1 : !mhlo.token
תכונות: AlwaysSpeculatableImplTrait
ממשקים: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
inputs | וריאדית של אסימון |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | אֲסִימוֹן |
mhlo.all_gather
(mhlo::AllGatherOp)
פעולת AllGather
בתוך כל קבוצת תהליך ברשת התהליך, משרשרת את הערכים של טנסור האופרנד מכל תהליך לאורך all_gather_dim
ומייצר טנזור תוצאה. computation
מיושם בנפרד עבור כל אופרנד operands
, ומפיק תוצאה אחת לכל אופרנד.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#all_gather
דוּגמָה:
%result = "mhlo.all_gather"(%operand) {
all_gather_dim = 1 : i64,
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>
// channel_id = 0
channel_handle = #mhlo.channel_handle<handle = 0, type = 0>,
// use_global_device_ids = false
} : (tensor<2x2xf32>) -> tensor<2x4xf32>
תכונות: SameOperandsAndResultElementType
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
all_gather_dim | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
replica_groups | ::mlir::DenseIntElementsAttr | תכונת רכיבי מספר שלם ללא סימן של 64 סיביות |
channel_handle | ::mlir::mhlo::ChannelHandleAttr | שני מספרים שלמים של 64 סיביות 'ידית' ו'סוג' |
use_global_device_ids | ::mlir::UnitAttr | תכונת יחידה |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operands | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
mhlo.all_reduce
(mhlo::AllReduceOp)
AllReduce פעולה
בתוך כל קבוצת תהליך ברשת התהליך, מיישמת computation
פונקציית הפחתה על הערכים של טנסור אופרנד מכל תהליך ומפיקה טנזור תוצאה. computation
מיושם בנפרד עבור כל אופרנד operands
, ומפיק תוצאה אחת לכל אופרנד.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#all_reduce
דוּגמָה:
%result = "mhlo.all_reduce"(%operand) ({
^bb0(%arg0: tensor<f32>, %arg1: tensor<f32>):
%0 = mhlo.add %arg1, %arg2 : tensor<f32>
mhlo.return %0 : tensor<f32>
}) {
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>
// channel_id = 0
channel_handle = #mhlo.channel_handle<handle = 0, type = 0>
// use_global_device_ids = false
} : (tensor<4xf32>) -> tensor<4xf32>
תכונות: InferTensorType
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
ממשקים: InferShapedTypeOpInterface
, InferTypeOpInterface
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
replica_groups | ::mlir::DenseIntElementsAttr | תכונת רכיבי מספר שלם ללא סימן של 64 סיביות |
channel_handle | ::mlir::mhlo::ChannelHandleAttr | שני מספרים שלמים של 64 סיביות 'ידית' ו'סוג' |
use_global_device_ids | ::mlir::UnitAttr | תכונת יחידה |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operands | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
mhlo.all_to_all
(mhlo::AllToAllOp)
פעולת AllToAll
בתוך כל קבוצת תהליכים ברשת התהליך, מפצל את ערכי הטנזור operand
לאורך split_dimension
לחלקים, מפזר את החלקים המפוצלים בין התהליכים, משרשר את החלקים המפוזרים לאורך concat_dimension
ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#all_to_all
דוּגמָה:
%result = "mhlo.all_to_all"(%operand) {
split_dimension = 1 : i64,
concat_dimension = 0 : i64,
split_count = 2 : i64,
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>
} : (tensor<2x4xf32>) -> tensor<4x2xf32>
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
, SameOperandsElementType
, SameOperandsShape
, SameVariadicOperandSize
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
split_dimension | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
concat_dimension | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
split_count | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שערכה חיובי |
replica_groups | ::mlir::DenseIntElementsAttr | תכונת רכיבי מספר שלם ללא סימן של 64 סיביות |
channel_handle | ::mlir::mhlo::ChannelHandleAttr | שני מספרים שלמים של 64 סיביות 'ידית' ו'סוג' |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכי מספר שלם ללא סימן |
mhlo.and
(mhlo::AndOp)
וגם תפעול
תַחבִּיר:
operation ::= `mhlo.and` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
מבצע AND מבחינה אלמנטית של שני טנסורים lhs
ו- rhs
ומייצר טנזור result
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#and
דוּגמָה:
%result = mhlo.and %lhs, %rhs : tensor<2x2xi32>
תכונות: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
lhs | טנסור מדורג של Pred (AKA בווליאני או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות חסר סימנים או ערכי מספר שלם ללא סימן 2/4/8/16/32/64 סיביות |
rhs | טנסור מדורג של Pred (AKA בווליאני או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות חסר סימנים או ערכי מספר שלם ללא סימן 2/4/8/16/32/64 סיביות |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16/32 סיביות אחידה עם סימן שלם או 2/4/8/16 /32-bit אחיד מכומד לכל מספר שלם בסימן או 2/4/8/16/32-bit אחיד מכומד לכל ציר ערכים שלמים ללא סימן |
mhlo.async_done
(mhlo::AsyncDoneOp)
פעולת AsyncDone
פעולה זו פרטית ל-XLA מהדר, ולכן אין לה עדיין מפרט.
באופן לא רשמי, פעולה זו חוסמת עד לסיום חישוב אסינכרוני. הוא מחזיר את התוצאה הסופית של החישוב האסינכרוני.
עיין בתיעוד עבור AsyncStart למידע נוסף.
ממשקים: InferTypeOpInterface
אופרנדים:
אופרנד | תֵאוּר |
---|---|
bundle | async_bundle עם כל שילוב של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E3M4 או סוג NU או 16-bit float או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit ללא סימנים או 2/4/8/16/32/64-bit ללא סימן מספר שלם או סוג מורכב עם רכיבים צפים של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם או 2/4/8/16/32 סיביות אחיד כמותי ללא סימן או 2/ 4/8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן ציר או 2/4/8/16/32-bit אחיד מכומד לפי ציר ערכי מספר שלם ללא סימן או אסימון או טופל מקונן עם כל שילוב של טנזור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או f6E3FN2M2 סוג או f8E3M4 או סוג f8E4M3 או סוג f8E4M3FN או סוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2FNUZ או סוג f8E8M0FNU או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M2FNUZ או מסוג f8E8M0FNU או 16-bit float או 32-bit float או KA float או 6-bit float 1 -bit שלם) או 2/4/8/16/32/64 סיביות חסר סימנים או 2/4/8/16/32/64 סיביות ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או 64 סיביות או 2/4/8/16/32 סיביות אחיד מרוכז בסימן שלם או 2/4/8/16/32 סיביות אחיד מקומת ללא סימן ערכי מספר שלם או טנזור מדורג של 2/4/8/16/32 סיביות אחיד מכומד מספר שלם בסימן לכל ציר או 2/4/8/16/32 סיביות אחיד בכימות לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
mhlo.async_start
(mhlo::AsyncStartOp)
פעולת AsyncStart
פעולה זו פרטית ל-XLA מהדר, ולכן אין לה עדיין מפרט.
באופן לא רשמי, פעולה זו מתחילה חישוב אסינכרוני.
זה משמש כאשר יש פונקציות המכילות גם המתנה אסינכרונית (כגון DMAs) וגם חישוב על-פתיל. לדוגמה, פונקציה עשויה להיות מורכבת מחישוב, DMA, חישוב אחר, DMA שני וחישוב סופי. זה יוצג כ-async_start ואחריו ו-async_update ו-async_done. ה-async_start יבצע את החישוב הראשון על השרשור ולאחר מכן יתחיל את ה-DMA. ה-async_update ימתין להשלמת ה-DMA אם הוא עדיין לא נעשה, ואז יבצע את החישוב השני בפונקציה, ויתחיל את ה-DMA השני. לבסוף, ה-async_done ימתין ב-DMA האחרון הזה, ולאחר מכן יריץ את החישוב האחרון שצריך להפעיל ב-thread ויחזיר את התוצאה של החישוב הסופי הזה.
operands
מועברים לחישוב ישירות called_computation
היא הפונקציה שתופעל באופן אסינכרוני execution_thread
הוא שם השרשור בו הוא יופעל. החוט הראשי נקרא "ראשי". לכל האשכולות יש שמות.
זה מחזיר את כל המצב הדרוש בין פעולות אסינכרון. לאחר הקצאת מאגר, ערכי ההחזרה מייצגים את השטח הדרוש כדי להחזיק את הקלט, התוצאות וכל משטחי הגירוד הדרושים או נערכו על ידי ה-async op.
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
called_computation | ::mlir::FlatSymbolRefAttr | תכונת הפניה לסמל שטוח |
execution_thread | ::mlir::StringAttr | תכונת מחרוזת |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
inputs | variadic של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3B11FNUZ או מסוג f8E52 או מסוג f8E52 או מסוג f8E52 -bit float או 32-bit float או 64- bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit מספר שלם ללא סימן או 2/4/8/16/32/64-bit ללא סימן מספר שלם או מורכב הקלד עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם קוונטי או 2/4/8/16/32 סיביות אחיד מכומתי ללא סימן או 2/4/8 /16/32-bit אחיד מכומד לכל מספר שלם בסימן ציר או 2/4/8/16/32-bit אחיד מכומד לפי ציר ערכי מספר שלם ללא סימן או אסימון או טופל מקונן עם כל שילוב של טנזור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או f6E3FN2M2 סוג או f8E3M4 או סוג f8E4M3 או סוג f8E4M3FN או סוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2FNUZ או סוג f8E8M0FNU או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M2FNUZ או מסוג f8E8M0FNU או 16-bit float או 32-bit float או KA float או 6-bit float 1 -bit שלם) או 2/4/8/16/32/64 סיביות חסר סימנים או 2/4/8/16/32/64 סיביות ללא סימן או סוג מורכב עם רכיבי float של 32 סיביות או 64 סיביות או 2/4/8/16/32 סיביות אחיד מרוכז בסימן שלם או 2/4/8/16/32 סיביות אחיד מקומת ללא סימן ערכי מספר שלם או טנזור מדורג של 2/4/8/16/32 סיביות אחיד מכומד מספר שלם בסימן לכל ציר או 2/4/8/16/32 סיביות אחיד בכימות לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | async_bundle עם כל שילוב של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E3M4 או סוג NU או 16-bit float או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit ללא סימנים או 2/4/8/16/32/64-bit ללא סימן מספר שלם או סוג מורכב עם רכיבים צפים של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם או 2/4/8/16/32 סיביות אחיד כמותי ללא סימן או 2/ 4/8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
mhlo.async_update
(mhlo::AsyncUpdateOp)
פעולת AsyncUpdate
פעולה זו פרטית ל-XLA מהדר, ולכן אין לה עדיין מפרט.
באופן לא רשמי, פעולה זו חוסמת חישוב אסינכרוני עד למחסום סנכרון. זה מחזיר bundle
לאחר הפעלתו.
עיין בתיעוד עבור AsyncStart למידע נוסף.
Interfaces: InferTypeOpInterface
אופרנדים:
אופרנד | תֵאוּר |
---|---|
bundle | async_bundle עם כל שילוב של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E3M4 או סוג NU או 16-bit float או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit ללא סימנים או 2/4/8/16/32/64-bit ללא סימן מספר שלם או סוג מורכב עם רכיבים צפים של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם או 2/4/8/16/32 סיביות אחיד כמותי ללא סימן או 2/ 4/8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
"ללא שם" | async_bundle עם כל שילוב של טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E3M4 או סוג NU או 16-bit float או 32-bit float או 64-bit float או bfloat16 סוג או pred (AKA בווליאני או 1-bit מספר שלם) או 2/4/8/16/32/64-bit ללא סימנים או 2/4/8/16/32/64-bit ללא סימן מספר שלם או סוג מורכב עם רכיבים צפים של 32 סיביות או רכיבי ציפה של 64 סיביות או 2/4/8/16/32 סיביות אחיד עם סימן שלם או 2/4/8/16/32 סיביות אחיד כמותי ללא סימן או 2/ 4/8/16/32 סיביות כמותיות אחידות לכל מספר שלם בסימן או 2/4/8/16/32 סיביות אחידה מכומדת לכל ציר ערכי מספר שלם ללא סימן או ערכי אסימון |
mhlo.atan2
(mhlo::Atan2Op)
מבצע Atan2
תַחבִּיר:
operation ::= `mhlo.atan2` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
מבצע פעולת atan2 מבחינת אלמנט על טנזור lhs
ו- rhs
ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#atan2
דוּגמָה:
%result = mhlo.atan2 %lhs, %rhs : tensor<3xf32>
תכונות: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
lhs | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או סוג bfloat16 או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או ערכי מספר שלם עם סימן שלם קוונטי אחיד של 2/4/8/16/32 סיביות או 2/4/8/16/32 סיביות אחידים ללא סימן |
rhs | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או סוג bfloat16 או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או ערכי מספר שלם עם סימן שלם קוונטי אחיד של 2/4/8/16/32 סיביות או 2/4/8/16/32 סיביות אחידים ללא סימן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או סוג bfloat16 או סוג מורכב עם רכיבי float של 32 סיביות או רכיבי float של 64 סיביות או ערכי מספר שלם עם סימן שלם קוונטי אחיד של 2/4/8/16/32 סיביות או 2/4/8/16/32 סיביות אחידים ללא סימן |
mhlo.batch_norm_grad
(mhlo::BatchNormGradOp)
פעולת BatchNormGrad
מחשב גרדיאנטים של מספר כניסות של BatchNormTrainingOp המתפשטות בחזרה מ- grad_output
, ומייצר טנסורים grad_operand
, grad_scale
ו- grad_offset
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#batch_norm_grad
דוּגמָה:
%grad_operand, %grad_scale, %grad_offset =
"mhlo.batch_norm_grad"(%operand, %scale, %mean, %variance, %grad_output) {
epsilon = 0.0 : f32,
feature_index = 2 : i64
} : (tensor<2x2x2xf32>, tensor<2xf32>, tensor<2xf32>, tensor<2xf32>,
tensor<2x2x2xf32>) -> (tensor<2x2x2xf32>, tensor<2xf32>, tensor<2xf32>)
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
epsilon | ::mlir::FloatAttr | תכונת ציפה של 32 סיביות |
feature_index | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
scale | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
mean | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
variance | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
grad_output | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
grad_operand | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
grad_scale | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
grad_offset | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
mhlo.batch_norm_inference
(mhlo::BatchNormInferenceOp)
פעולת BatchNormInference
מנרמל את טנסור operand
על פני כל הממדים מלבד הממד feature_index
ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#batch_norm_inference
דוּגמָה:
%result = "mhlo.batch_norm_inference"(%operand, %scale, %offset, %mean, %variance) {
epsilon = 0.0 : f32,
feature_index = 2 : i64
} : (tensor<2x2x2xf32>, tensor<2xf32>, tensor<2xf32>, tensor<2xf32>, tensor<2xf32>) -> tensor<2x2x2xf32>
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
epsilon | ::mlir::FloatAttr | תכונת ציפה של 32 סיביות |
feature_index | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
scale | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
offset | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
mean | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
variance | טנסור 1D מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או מסוג f8E5M5M2 או מסוג f8E5M5E או f8E5M2 או מסוג f8F8E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג f4E2M1FN או מסוג f6E2M3FN או מסוג f6E3M2FN או מסוג f8E3M4 או מסוג f8E4M3 או מסוג f8E4M3FN או מסוג f8E4M3FNUZ או מסוג f8E4M3B11FNUZ או סוג f8E5M2 או מסוג f8E5M2 או f8E5M2 או מסוג f8E4E או 32-bit float או 64-bit float או ערכי סוג bfloat16 |
mhlo.batch_norm_training
(mhlo::BatchNormTrainingOp)
פעולת BatchNormTraining
מחשב את הממוצע והשונות על פני מימדים אצווה ומרחביים ומנרמל את טנסור operand
, עבור כל תכונה בממד feature_index
ומייצר טנסור output
, batch_mean
ו- batch_var
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#batch_norm_training
דוּגמָה:
%output, %batch_mean, %batch_var = "mhlo.batch_norm_training"(%operand, %scale, %offset) {
epsilon = 0.0 : f32,
feature_index = 2 : i64
} : (tensor<2x2x2xf32>, tensor<2xf32>, tensor<2xf32>) -> (tensor<2x2x2xf32>, tensor<2xf32>, tensor<2xf32>)
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
epsilon | ::mlir::FloatAttr | תכונת ציפה של 32 סיביות |
feature_index | ::mlir::IntegerAttr | תכונת מספר שלם ללא סימן של 64 סיביות שהערך שלה אינו שלילי |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
scale | טנסור 1D מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
offset | טנסור 1D מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
batch_mean | טנסור 1D מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
batch_var | טנסור 1D מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או ערכי סוג BFLOAT16 |
mhlo.bitcast
(mhlo :: bitcastop)
פעולת Bitcast
תַחבִּיר:
operation ::= `mhlo.bitcast` operands attr-dict `:` functional-type(operands, results)
פעולה זו היא פרטית למהדר XLA, ולכן אין לה עדיין מפרט.
באופן לא פורמלי, פעולה זו משנה את צורת הקלט באופן בו הסידור הפיזי של האלמנטים אינו משתנה.
פעולה זו זקוקה למידע פריסה כדי להבין את "סידור פיזי של אלמנטים", ותמיכה בפריסה ב- MHLO היא כיום יצירה שמתבצעת.
דוּגמָה:
%0 = mhlo.bitcast %arg0 : (tensor<3x4xf32>) -> tensor<3x4x1xf32>
תכונות: AlwaysSpeculatableImplTrait
ממשקים: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.bitcast_convert
(mhlo :: bitcastconvertop)
פעולת BitcastConvert
תַחבִּיר:
operation ::= `mhlo.bitcast_convert` operands attr-dict `:` functional-type(operands, results)
מבצע פעולת טיסה ב- operand
Tensor ומייצר טנזור result
בו מתפרשים חתיכות של Tensor operand
כולו באמצעות סוג Tensor result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#bitcast_convert
דוּגמָה:
%result = mhlo.bitcast_convert %operand : (tensor<2xf32>) -> tensor<2x4xi8>
תכונות: AlwaysSpeculatableImplTrait
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.broadcast
(mhlo :: broadcastop)
פעולת שידור
פעולה זו בדרך החוצה מ- StableHlo, כך שהיא אינה כלולה במפרט: https://github.com/openxla/stablehlo/issues/3
באופן לא פורמלי, פעולה זו עושה את אותו הדבר כמו השידור של XLA: https://www.tensorflow.org/xla/operation_semantics#broadcast
דוּגמָה:
%result = mhlo.broadcast %operand, sizes = [1, 2] : (tensor<3xi32>) -> tensor<1x2x3xi32>
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
, SameOperandsAndResultElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
broadcast_sizes | :: mlir :: densintelementsattr | תכונה של תכונה של מספר שלם שלם 64 סיביות |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.broadcast_in_dim
(mhlo :: broadcastindimop)
פעולת BroadcastIndim
מרחיב את הממדים ו/או דרגת טנזור קלט על ידי שכפול הנתונים במתחם operand
ומייצר טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#broadcast_in_dim
דוּגמָה:
%result = mhlo.broadcast_in_dim %operand, dims = [2, 1] : (tensor<1x3xi32>) -> tensor<2x3x2xi32>
תכונות: AlwaysSpeculatableImplTrait
HLO_CompatibleOperandsAndResultElementType
ממשקים: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
broadcast_dimensions | :: mlir :: densintelementsattr | תכונה של תכונה של מספר שלם שלם 64 סיביות |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מעוצב סטטי מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E5M2 STORE או F8E4PNUZ סוג F8E5M2M2M2 צף סיביות או צף 32 סיביות או 64 סיביות float או bfloat16 סוג או pred (aka boolean או מספר שלם 1-bit) או 2/4/8/16/32/64 סיב עם ציפה של 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/8 כמות כמותית של 16/32 סיביות כמותית לכל ציר חתום שלם או 2/4/8/16/32 סיב |
mhlo.case
(mhlo :: caseop)
פעולת מקרה
מייצר את הפלט מביצוע function
אחת בדיוק branches
בהתאם לערך index
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#case
דוּגמָה:
%result0, %result1 = "mhlo.case"(%index) ({
mhlo.return %result_branch0, %result_branch0 : tensor<2xi64>, tensor<2xi64>
}, {
mhlo.return %result_branch1, %result_branch1 : tensor<2xi64>, tensor<2xi64>
}) : (tensor<i32>) -> (tensor<2xi64>, tensor<2xi64>)
תכונות: RecursiveMemoryEffects
, SingleBlockImplicitTerminator<ReturnOp>
SingleBlock
ממשקים: InferTypeOpInterface
אופרנדים:
אופרנד | תֵאוּר |
---|---|
index | טנזור של ערכי מספר שלם נטולי אותות של 32 סיביות |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | משתנה של טנזור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 או F8E4M3FN סוג או F8E42M2MOS או F8E5m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2222 -הביט צף או צף 32 סיביות או 64- ביט צף או bfloat16 סוג או pred (aka Boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיב הקלד עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים כמותים חתומים או 2/4/8/16/32 סיביות אחידים ערכי שלם לא חתומים או טנזור מדורגים של 2 2 /4/8/16/32 סיביות אחידות כמותיות לכל ציר חתום שלם שלם או 2/4/8/16/32 סיב |
mhlo.cbrt
(mhlo :: cbrtop)
פעולת CBRT
תַחבִּיר:
operation ::= `mhlo.cbrt` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
מבצע פעולת שורש מעוקבת בחינת אלמנט על טנזור operand
ומייצר טנזור result
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#cbrt
דוּגמָה:
%result = mhlo.cbrt %operand : tensor<4xf32>
תכונות: AlwaysSpeculatableImplTrait
Elementwise
, CompatibleOperandsAndResultType
SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או סוג Bfloat16 או סוג מורכב עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או סוג Bfloat16 או סוג מורכב עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיב |
mhlo.ceil
(mhlo :: ceilop)
פעולת תקרת
תַחבִּיר:
operation ::= `mhlo.ceil` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
מבצע תקרת אלמנטים של אלמנט של operand
טנזור ומייצר טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#ceil
דוּגמָה:
%result = mhlo.ceil %operand : tensor<5xf32>
תכונות: AlwaysSpeculatableImplTrait
Elementwise
, CompatibleOperandsAndResultType
SameOperandsAndResultShape
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או 2/4/8/16/32 סיביות אחידות כמותיות חתומות כמותיות או 2/4/8/16/32 סיביות אחידות ערכי שלם לא חתומים |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או 2/4/8/16/32 סיביות אחידות כמותיות חתומות כמותיות או 2/4/8/16/32 סיביות אחידות ערכי שלם לא חתומים |
mhlo.cholesky
(mhlo :: choleskyop)
מבצע צ'ולסקי
מחשב את הפירוק של צ'ולסקי של חבורה של מטריצות.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#cholesky
דוּגמָה:
%result = mhlo.cholesky %a, lower = true : tensor<3x3xf32>
תכונות: AlwaysSpeculatableImplTrait
, InferTensorType
, SameOperandsAndResultElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
lower | :: mlir :: boolattr | תכונה Bool |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
a | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או סוג BFLOAT16 או סוג מורכב עם ערכי ציפה של 32 סיביות או ערכי אלמנטים צפים של 64 סיביות |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או סוג BFLOAT16 או סוג מורכב עם ערכי ציפה של 32 סיביות או ערכי אלמנטים צפים של 64 סיביות |
mhlo.clamp
(mhlo :: clampop)
פעולת מהדק
תַחבִּיר:
operation ::= `mhlo.clamp` $min `,` $operand `,` $max attr-dict
`:` custom<SameOperandsAndResultType>(type($min), type($operand), type($max), type($result))
מהדק כל אלמנט במתחם operand
בין ערך מינימלי למקסימום ומייצר טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#clamp
דוּגמָה:
%result = mhlo.clamp %min, %operand, %max : tensor<3xi32>
תכונות: AlwaysSpeculatableImplTrait
speculatableimpltrait, HLO_BroadcastingElementwise
, InferTensorType
, SameOperandsAndResultElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
min | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
max | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.collective_broadcast
(mhlo :: collectivebroadcastop)
פעולת קולקטיביות משודרת
בתוך כל קבוצת תהליכים ברשת התהליך, שלח את הערך של Tendor operand
מתהליך המקור לתהליכי היעד והפיק טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#collective_broadcast
דוּגמָה:
%result = "mhlo.collective_broadcast"(%operand) {
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>,
channel_handle = #mhlo.channel_handle<handle = 0, type = 0>
} : (tensor<1x2xi64>) -> tensor<1x2xi64>
תכונות: CompatibleOperandsAndResultType
ממשקים: InferShapedTypeOpInterface
, InferTypeOpInterface
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
replica_groups | :: mlir :: densintelementsattr | תכונה של תכונה של מספר שלם שלם 64 סיביות |
channel_handle | :: mlir :: mhlo :: ChannelHandLeattr | שני מספרים שלמים של 64 סיביות 'ידית' ו- 'סוג' |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.collective_permute
(mhlo :: oldivepermuteop)
פעולה קולקטיבית
בתוך כל קבוצת תהליכים ברשת התהליך, שולח את הערך של Tensor operand
מתהליך המקור לתהליך היעד ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#collective_permute
דוּגמָה:
%result = "mhlo.collective_permute"(%operand) {
source_target_pairs = dense<[[0, 1], [1, 2]]> : tensor<2x2xi64>,
// channel_id = 0
channel_handle = #mhlo.channel_handle<handle = 0, type = 0>
} : (tensor<4x2xf32>) -> tensor<4x2xf32>
תכונות: AlwaysSpeculatableImplTrait
CompatibleOperandsAndResultType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
source_target_pairs | :: mlir :: densintelementsattr | תכונה של תכונה של מספר שלם שלם 64 סיביות |
channel_handle | :: mlir :: mhlo :: ChannelHandLeattr | שני מספרים שלמים של 64 סיביות 'ידית' ו- 'סוג' |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
operand | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.compare
(mhlo :: compareop)
השווה פעולה
תַחבִּיר:
operation ::= `mhlo.compare` $comparison_direction `,` $lhs `,` $rhs (`,` $compare_type^)?
attr-dict `:` functional-type(operands, results)
מבצע השוואה בין אלמנטים של טנזורי lhs
ו- rhs
בהתאם comparison_direction
והשוואה_ compare_type
, ומייצר טנזור result
.
ראה: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#compare
דוּגמָה:
%result = mhlo.compare LT, %lhs, %rhs, FLOAT : (tensor<2xf32>, tensor<2xf32>) -> tensor<2xi1>
תכונות: AlwaysSpeculatableImplTrait
, Elementwise
, InferTensorType
, SameOperandsAndResultShape
, SameOperandsElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
comparison_direction | :: mlir :: mhlo :: השוואת DirectionAttr | איזו פעולת השוואה לביצוע. |
compare_type | :: mlir :: mhlo :: השוואה typeattr | באיזה סוג השוואה יש להשתמש. |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
lhs | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
rhs | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנזור מדורג של ערכים של מספר שלם בוליאני או 1 סיביות) |
mhlo.complex
(mhlo :: complexop)
פעולה מורכבת
תַחבִּיר:
operation ::= `mhlo.complex` operands attr-dict
`:` custom<ComplexOpType>(type($lhs), type($rhs), type($result))
מבצע המרה חכמת אלמנט לערך מורכב מזוג ערכים אמיתיים ודמיוניים, lhs
ו- rhs
, ומייצר טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#complex
דוּגמָה:
%result = mhlo.complex %lhs, %rhs : tensor<2xcomplex<f32>>
תכונות: AlwaysSpeculatableImplTrait
SpeculatableImpltrait, Elementwise
, SameOperandsAndResultShape
, SameOperandsElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
אופרנדים:
אופרנד | תֵאוּר |
---|---|
lhs | טנזור מדורג של צף 32 סיביות או ערכי ציפה של 64 סיביות |
rhs | טנזור מדורג של צף 32 סיביות או ערכי ציפה של 64 סיביות |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | טנזור מדורג מסוג מורכב עם צף 32 סיביות או ערכי אלמנטים צפים של 64 סיביות |
mhlo.composite
(mhlo :: compositeop)
פעולה מורכבת
תַחבִּיר:
operation ::= `mhlo.composite` $name $inputs attr-dict `:` functional-type(operands, results)
עוטף פעולה המורכבת (מורכבת) של פעולות אחרות של StavelHlo, נטילת inputs
composite_attributes
ומייצרת results
. הסמנטיקה של ה- OP מיושמת על ידי תכונת decomposition
. ניתן להחליף את ה- OP composite
בפירוק שלה מבלי לשנות סמנטיקה של התוכנית. במקרים שבהם הפירוק הפירוק אינו מספק את אותה סמנטיקה של OP, העדיפו להשתמש custom_call
.
שדה version
(ברירת המחדל ל- 0
) משמש לציון כאשר הסמנטיקה של המורכב משתנה.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#composite
דוּגמָה:
%results = mhlo.composite "my.op" %arg0, %arg1 {
decomposition = @my_op,
composite_attributes = { my_attribute = "my_value" },
version = 1 : i32
} : (tensor<f32>, tensor<f32>) -> tensor<f32>
ממשקים: SymbolUserOpInterface
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
name | :: mlir :: stringattr | תכונה מחרוזת |
composite_attributes | :: mlir :: Dictionaryattr | מילון ערכי תכונה ששמו |
decomposition | :: mlir :: flatsymbolrefattr | תכונה התייחסות לסמל שטוח |
version | :: mlir :: integerattr | תכונה שלם שלם ללא אותות 32 סיביות |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
inputs | משתנה של טנזור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 או F8E4M3FN סוג או F8E42M2MOS או F8E5m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2222 -הביט צף או צף 32 סיביות או 64- ביט צף או bfloat16 סוג או pred (aka Boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיב הקלד עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיב /16/32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיביות אחידות כמותיות לציר ערכי מספר שלם לא חתום או אסימון או טופל מקונן עם כל שילוב של טנסור מדורג מסוג F4E2M1FN או F6E2M3FN סוג או F6E3M2FN סוג סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ סוג או F8E5M2 סוג או F8E5M2FNUZ סוג או F8E8M0FNU סוג או 16-BIT OR 1-BIT OR OR OR OR OR OR (APPLOAT OR OR, OR OR OR OR OR OR OR OR OR OR OR OR OR OR או או BIT. -ל שלם שלם) או 2/4/8/16/32/64 סיביות מספר שלם שלם או 2/4/8/16/32/64 סיב או 2/4/8/16/32 סיביות אחידות כמותיות שלם חתום או 2/4/8/16/32 סיב לכל ציר חתום שלם שלם או 2/4/8/16/32 סיביות כמותית כמותית לכל ציר ערכי מספר שלם או ערכי אסימון לא חתומים או ערכי אסימון |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | משתנה של טנזור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 או F8E4M3FN סוג או F8E42M2MOS או F8E5m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2222 -הביט צף או צף 32 סיביות או 64- ביט צף או bfloat16 סוג או pred (aka Boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיב הקלד עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיב /16/32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיביות אחידות כמותיות לציר ערכי מספר שלם לא חתום או אסימון או טופל מקונן עם כל שילוב של טנסור מדורג מסוג F4E2M1FN או F6E2M3FN סוג או F6E3M2FN סוג סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ סוג או F8E5M2 סוג או F8E5M2FNUZ סוג או F8E8M0FNU סוג או 16-BIT OR 1-BIT OR OR OR OR OR OR (APPLOAT OR OR, OR OR OR OR OR OR OR OR OR OR OR OR OR OR או או BIT. -ל שלם שלם) או 2/4/8/16/32/64 סיביות מספר שלם שלם או 2/4/8/16/32/64 סיב או 2/4/8/16/32 סיביות אחידות כמותיות שלם חתום או 2/4/8/16/32 סיב לכל ציר חתום שלם שלם או 2/4/8/16/32 סיביות כמותית כמותית לכל ציר ערכי מספר שלם או ערכי אסימון לא חתומים או ערכי אסימון |
mhlo.concatenate
(mhlo :: concatenateop)
פעולת שרשור
משרשר מספר משתנה של טנזורים inputs
לאורך ממד dimension
באותו סדר כמו הטיעונים הנתונים ומייצר טנזור result
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#concatenate
דוּגמָה:
%result = mhlo.concatenate %input0, %input1, dim = 0 : (tensor<3x2xi64>, tensor<1x2xi64>) -> tensor<4x2xi64>
תכונות: AlwaysSpeculatableImplTrait
SameOperandsAndResultElementType
ממשקים: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
אפקטים: MemoryEffects::Effect{}
תכונות:
תְכוּנָה | סוג MLIR | תֵאוּר |
---|---|---|
dimension | ::mlir::IntegerAttr | תכונה שלם שלם של מספר שלם 64 סיביות שערכתה אינה שלילית |
אופרנדים:
אופרנד | תֵאוּר |
---|---|
val | משתנה של טנזור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 או F8E4M3FN סוג או F8E42M2MOS או F8E5m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2m2222 -הביט צף או צף 32 סיביות או 64- ביט צף או bfloat16 סוג או pred (aka Boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיב הקלד עם צף 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיב /16/32 סיביות אחידות כמותיות לכל ציר חתום שלם שלם או 2/4/8/16/32 סיב |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«ללא שם» | טנסור מדורג מסוג F4E2M1FN סוג או F6E2M3FN סוג או F6E3M2FN סוג או F8E3M4 סוג או F8E4M3 סוג או F8E4M3FN סוג או F8E4M3FNUZ סוג או F8E4M3B11FNUZ או F8E5M2 OR או F8E5MUZ או צף 32 סיביות או צף 64 סיביות או bfloat16 סוג או pred (aka boolean או מספר שלם של 1 סיביות) או 2/4/8/16/32/64 סיביות מספר שלם או 2/4/8/16/32/64 סיב ציפה 32 סיביות או אלמנטים צפים של 64 סיביות או 2/4/8/16/32 סיביות אחידים כמותים שלם חתום או 2/4/8/16/32 סיביות אחידים כמותים לא מסותמים או 2/4/8/16 אחידים. /32 סיביות אחידות כמותיות לכל ציר חתום מספר שלם או 2/4/8/16/32 סיב |
mhlo.constant
(mhlo :: constantop)
פעולה מתמדת
מייצר output
פלט מערך value
.
ראו: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#constant
דוּגמָה:
%output = mhlo.constant dense<[[0.0, 1.0], [2.0, 3.0]]> : tensor<2x2xf32>
תכונות: AlwaysSpeculatableImplTrait
, ConstantLike
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
value | ::mlir::ElementsAttr | constant vector/tensor attribute |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.convert
(mhlo::ConvertOp)
Convert operation
תַחבִּיר:
operation ::= `mhlo.convert` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs an element-wise conversion from one element type to another on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#convert
דוּגמָה:
%result = mhlo.convert %operand : (tensor<3xi32>) -> tensor<3xcomplex<f32>>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.convolution
(mhlo::ConvolutionOp)
Convolution operation
תַחבִּיר:
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 dot products between windows of lhs
and slices of rhs
and produces result
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#convolution
דוּגמָה:
%result = "mhlo.convolution"(%lhs, %rhs) {
window_strides = dense<4> : tensor<2xi64>,
padding = dense<0> : tensor<2x2xi64>,
lhs_dilation = dense<2> : tensor<2xi64>,
rhs_dilation = dense<1> : tensor<2xi64>,
window_reversal = dense<false> : tensor<2xi1>,
dimension_numbers = #mhlo.conv<[b, 0, 1, f]x[0, 1, i, o]->[b, 0, 1, f]>,
feature_group_count = 1 : i64,
batch_group_count = 1 : i64,
precision_config = [#stablehlo<precision DEFAULT>, #stablehlo<precision DEFAULT>]
} : (tensor<1x4x4x1xi32>, tensor<3x3x1x1xi32>) -> tensor<1x2x2x1xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 whose value is positive |
batch_group_count | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is positive |
precision_config | ::mlir::ArrayAttr | Precision Config attribute |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.copy
(mhlo::CopyOp)
Copy operation
תַחבִּיר:
operation ::= `mhlo.copy` operands attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
This operation is private to the XLA compiler, so it is does not yet have a specification.
Informally, this operation a copy of operand
. Depending on the metadata attached to the operation, it can behave quite differently from a no-op.
דוּגמָה:
%0 = mhlo.copy %arg0 : tensor<f32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
cross_program_prefetch_index | ::mlir::IntegerAttr | 32-bit signless integer attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
mhlo.cosine
(mhlo::CosineOp)
Cosine operation
תַחבִּיר:
operation ::= `mhlo.cosine` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise cosine operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#cosine
דוּגמָה:
%result = mhlo.cosine %operand : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.count_leading_zeros
(mhlo::ClzOp)
Clz operation
תַחבִּיר:
operation ::= `mhlo.count_leading_zeros` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise count of the number of leading zero bits in the operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#count_leading_zeros
דוּגמָה:
%result = mhlo.count_leading_zeros %operand : tensor<2x2xi8>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.create_token
(mhlo::CreateTokenOp)
CreateToken operation
תַחבִּיר:
operation ::= `mhlo.create_token` attr-dict `:` type(results)
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/3
Informally, this operation does the same thing as AfterAllOp with 0 inputs: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#after_all
דוּגמָה:
%output = mhlo.create_token : !mhlo.token
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | אֲסִימוֹן |
mhlo.cross-replica-sum
(mhlo::CrossReplicaSumOp)
CrossReplicaSum operation
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/3
Informally, this operation does the same thing as AllReduceOp with channel_id = 0
, use_global_device_ids = false
and computation
implementing addition: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#all_reduce
דוּגמָה:
%result = "mhlo.cross-replica-sum"(%operand) {
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>
} : (tensor<4xf32>) -> tensor<4xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
replica_groups | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.custom_call
(mhlo::CustomCallOp)
CustomCall operation
תַחבִּיר:
operation ::= `mhlo.custom_call` custom<CustomCallTarget>($call_target_name) `(` $inputs `)`
attr-dict `:` functional-type(operands, results)
Encapsulates an implementation-defined operation call_target_name
that takes inputs
and called_computations
and produces results
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#custom_call
דוּגמָה:
%results = "mhlo.custom_call"(%input0) {
call_target_name = "foo",
has_side_effect = false,
backend_config = "bar",
api_version = 1 : i32,
called_computations = [@foo]
} : (tensor<f32>) -> tensor<f32>
A custom call invokes code external to XLA. The `inputs` 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`.
If XLA runtime is enabled for a backend, then custom calls use the runtime
custom call calling convention to call into the external functions. This
calling convention defines an ABI for encoding arguments, attributes and
results.
Depending on the API version there are two ways to pass extra bits of static
information to the external function:
1. For `API_VERSION_TYPED_FFI` custom calls `backend_config` must be a
dictionary attribute, that will be encoded according to the custom call
calling convention and passed to the external function as the attributes
argument. External code is expected to use declarative bindings (see
`xla/runtime/custom_call.h`) to decode them at run time. These custom
calls are only supported if XLA uses XLA runtime.
2. For previous API versions it is the user responsibility to encode extra
bits of static information as a string `backend_config` attribute, and
decode it at run time.
Interfaces: MemoryEffectOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
call_target_name | ::mlir::StringAttr | string attribute |
has_side_effect | ::mlir::BoolAttr | bool attribute |
backend_config | ::mlir::Attribute | string attribute or dictionary of named attribute values |
api_version | ::mlir::mhlo::CustomCallApiVersionAttr | Custom call API version |
called_computations | ::mlir::ArrayAttr | flat symbol ref array attribute |
custom_call_schedule | ::mlir::mhlo::CustomCallScheduleAttr | Specifies the desired schedule for the custom-call. |
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 | תֵאוּר |
---|---|
inputs | variadic of tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
mhlo.divide
(mhlo::DivOp)
Div operation
תַחבִּיר:
operation ::= `mhlo.divide` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise division of dividend lhs
and divisor rhs
tensors and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#divide
דוּגמָה:
%result = mhlo.divide %lhs, %rhs : tensor<4xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.domain
(mhlo::DomainOp)
Domain operation
This operation is private to the XLA compiler, so it is does not yet have a specification.
Informally, these operations 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: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.dot
(mhlo::DotOp)
Dot operation
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/3
Informally, this operation does the same thing as XLA's Dot: https://www.tensorflow.org/xla/operation_semantics#dot
דוּגמָה:
%0 = mhlo.dot %arg0, %arg1 : (tensor<1x2xi32>, tensor<2x1xi32>) -> tensor<1x1xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
precision_config | ::mlir::ArrayAttr | Precision Config attribute |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dot_general
(mhlo::DotGeneralOp)
DotGeneral operation
Computes dot products between slices of lhs
and slices of rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dot_general
דוּגמָה:
%result = "mhlo.dot_general"(%lhs, %rhs) {
dot_dimension_numbers = #mhlo.dot<
lhs_batching_dimensions = [0],
rhs_batching_dimensions = [0],
lhs_contracting_dimensions = [2],
rhs_contracting_dimensions = [1]
>,
precision_config = [#stablehlo<precision DEFAULT>, #stablehlo<precision DEFAULT>]
} : (tensor<2x2x2xi32>, tensor<2x2x2xi32>) -> tensor<2x2x2xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dot_dimension_numbers | ::mlir::mhlo::DotDimensionNumbersAttr | Attribute that models the dimension information for dot. |
precision_config | ::mlir::ArrayAttr | Precision Config attribute |
algorithm | ::mlir::mhlo::DotAlgorithmAttr | Attribute that models the algorithm constraints to use for computing dot. |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_broadcast_in_dim
(mhlo::DynamicBroadcastInDimOp)
DynamicBroadcastInDim operation
This operation is functionally identical to broadcast_in_dim op, but the result shape is specified dynamically via output_dimensions
.
It also 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.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dynamic_broadcast_in_dim
דוּגמָה:
%operand = mhlo.constant dense<[[1, 2, 3]]> : tensor<1x3xi64>
%output_dimensions = mhlo.constant dense<[2, 3, 2]> : tensor<3xi64>
%result = "mhlo.dynamic_broadcast_in_dim"(%operand, %output_dimensions) {
broadcast_dimensions = array<i64: 2, 1>,
known_expanding_dimensions = array<i64: 0>,
known_nonexpanding_dimensions = array<i64: 1>
} : (tensor<1x3xi64>, tensor<3xi64>) -> tensor<2x3x2xi64>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
output_dimensions | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_conv
(mhlo::DynamicConvOp)
DynamicConv operation
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/8
Informally, this operation does the same thing as ConvolutionOp except that padding
is specified dynamically via d_padding
: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#convolution
דוּגמָה:
%result = "mhlo.dynamic_conv"(%lhs, %rhs, %d_padding) {
window_strides = dense<4> : tensor<2xi64>,
lhs_dilation = dense<2> : tensor<2xi64>,
rhs_dilation = dense<1> : tensor<2xi64>,
window_reversal = dense<false> : tensor<2xi1>,
dimension_numbers = #mhlo.conv<[b, 0, 1, f]x[0, 1, i, o]->[b, 0, 1, f]>,
feature_group_count = 1 : i64,
batch_group_count = 1 : i64,
precision_config = [#stablehlo<precision DEFAULT>, #stablehlo<precision DEFAULT>]
} : (tensor<1x4x4x1xi32>, tensor<3x3x1x1xi32>, tensor<2x2xi64>) -> tensor<1x2x2x1xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 whose value is positive |
batch_group_count | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is positive |
precision_config | ::mlir::ArrayAttr | Precision Config attribute |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
d_padding | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_gather
(mhlo::DynamicGatherOp)
DynamicGather operation
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/8
Informally, this operation does the same thing as GatherOp except that slice_sizes
are specified dynamically: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#gather
דוּגמָה:
%result = "mhlo.dynamic_gather"(%operand, %start_indices, %slice_sizes) {
dimension_numbers = #mhlo.gather<
offset_dims = [2, 3],
collapsed_slice_dims = [0],
start_index_map = [0, 2],
index_vector_dim = 2>,
indices_are_sorted = false
} : (tensor<3x4x2xi32>, tensor<2x3x2xi64>, tensor<3xi64>) -> tensor<2x3x2x2xi32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimension_numbers | ::mlir::mhlo::GatherDimensionNumbersAttr | Attribute that models the dimension information for gather |
indices_are_sorted | ::mlir::BoolAttr | bool attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
start_indices | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
slice_sizes | statically shaped 1-dimensional integer tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_iota
(mhlo::DynamicIotaOp)
DynamicIota operation
This operation is functionally identical to iota op, but the result shape is specified dynamically via output_shape
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dynamic_iota
דוּגמָה:
%0 = mhlo.dynamic_iota %arg0, dim = 0 : (tensor<1xindex>) -> tensor<4xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
iota_dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is non-negative |
Operands:
Operand | תֵאוּר |
---|---|
output_shape | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_pad
(mhlo::DynamicPadOp)
DynamicPad operation
תַחבִּיר:
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.
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
padding_value | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
edge_padding_low | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
edge_padding_high | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
interior_padding | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_reshape
(mhlo::DynamicReshapeOp)
DynamicReshape operation
תַחבִּיר:
operation ::= `mhlo.dynamic_reshape` operands attr-dict `:` functional-type(operands, results)
This operation is functionally identical to reshape op, but the result shape is specified dynamically via output_shape
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dynamic_reshape
דוּגמָה:
%output_shape = mhlo.constant dense<[3, 2]> : tensor<2xi64>
%result = mhlo.dynamic_reshape %operand, %output_shape : (tensor<2x3xi64>, tensor<2xi64>) -> tensor<3x2xi64>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
output_shape | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_slice
(mhlo::DynamicSliceOp)
DynamicSlice operation
Extracts a slice from the operand
using dynamically-computed starting indices and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dynamic_slice
דוּגמָה:
%result = mhlo.dynamic_slice %operand, %start_indices0, %start_indices1, sizes = [2, 2]
: (tensor<4x4xi32>, tensor<i64>, tensor<i64>) -> tensor<2x2xi32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
slice_sizes | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
start_indices | variadic of 0D tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.dynamic_update_slice
(mhlo::DynamicUpdateSliceOp)
DynamicUpdateSlice operation
תַחבִּיר:
operation ::= `mhlo.dynamic_update_slice` operands attr-dict `:` functional-type(operands, results)
Produces a result
tensor which is equal to the operand
tensor except that the slice starting at start_indices
is updated with the values in update
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#dynamic_update_slice
דוּגמָה:
%result = mhlo.dynamic_update_slice %operand, %update, %start_indices0, %start_indices1
: (tensor<4x4xi32>, tensor<2x2xi32>, tensor<i64>, tensor<i64>) -> tensor<4x4xi32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
update | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
start_indices | variadic of 0D tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.einsum
(mhlo::EinsumOp)
Einsum operation
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/3
Informally, this operation does the same thing as TF's einsum: https://www.tensorflow.org/api_docs/python/tf/einsum
דוּגמָה:
%result = "mhlo.einsum"(%lhs, %rhs) {
einsum_config = "ab,bc->ac"
} : (tensor<4x16xf32>, tensor<16x4xf32>) -> tensor<4x4xf32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
einsum_config | ::mlir::StringAttr | string attribute |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.erf
(mhlo::ErfOp)
Erf operation
תַחבִּיר:
operation ::= `mhlo.erf` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise erf operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#erf
דוּגמָה:
%result = mhlo.erf %operand : tensor<2x2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.exponential
(mhlo::ExpOp)
Exp operation
תַחבִּיר:
operation ::= `mhlo.exponential` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise exponential operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#exponential
דוּגמָה:
%result = mhlo.exponential %operand : tensor<2x2xf64>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.exponential_minus_one
(mhlo::Expm1Op)
Expm1 operation
תַחבִּיר:
operation ::= `mhlo.exponential_minus_one` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise exponential minus one operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#exponential_minus_one
דוּגמָה:
%result = mhlo.exponential_minus_one %operand : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.fft
(mhlo::FftOp)
Fft operation
Performs the forward and inverse Fourier transforms for real and complex inputs/outputs.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#fft
דוּגמָה:
%result = mhlo.fft %operand, type = FFT, length = [4] : (tensor<4xcomplex<f32>>) -> tensor<4xcomplex<f32>>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
fft_type | ::mlir::mhlo::FftTypeAttr | XLA fast fourier transform type. |
fft_length | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.floor
(mhlo::FloorOp)
Floor operation
תַחבִּיר:
operation ::= `mhlo.floor` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise floor of operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#floor
דוּגמָה:
%result = mhlo.floor %operand : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.fusion
(mhlo::FusionOp)
Fusion operation
This operation is private to the XLA compiler, so it is does not yet have a specification.
Informally, this operation 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:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
fusion_kind | ::mlir::mhlo::FusionKindAttr | fusion kind |
output_operand_aliases | ::mlir::ArrayAttr | Aliasing attribute for outputs and operands of Fusion |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
results | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
mhlo.gather
(mhlo::GatherOp)
Gather operation
Gathers slices from operand
tensor from offsets specified in start_indices
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#gather
דוּגמָה:
%result = "mhlo.gather"(%operand, %start_indices) {
dimension_numbers = #stablehlo.gather<
offset_dims = [3, 4],
collapsed_slice_dims = [1],
operand_batching_dims = [0],
start_indices_batching_dims = [1],
start_index_map = [2, 1],
index_vector_dim = 3>,
slice_sizes = dense<[0, 2, 2]> : tensor<3xi64>,
indices_are_sorted = false
} : (tensor<2x3x4x2xi64>, tensor<2x2x3x2xi64>) -> tensor<2x2x3x2x2xi64>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
start_indices | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.get_dimension_size
(mhlo::GetDimensionSizeOp)
GetDimensionSize operation
Produces the size of the given dimension
of the operand
.
See https://github.com/openxla/stablehlo/blob/main/docs/spec.md#get_dimension_size
דוּגמָה:
%result = mhlo.get_dimension_size %operand, dim = 1 : (tensor<2x3xf32>) -> tensor<i32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is non-negative |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | tensor of 32-bit signless integer values |
mhlo.get_tuple_element
(mhlo::GetTupleElementOp)
GetTupleElement operation
תַחבִּיר:
operation ::= `mhlo.get_tuple_element` $operand `[` $index `]` attr-dict `:` functional-type(operands, results)
Extracts element at index
position of the operand
tuple and produces a result
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#get_tuple_element
דוּגמָה:
%result = mhlo.get_tuple_element %operand[0] : (tuple<tensor<2xf32>, tuple<tensor<i32>>>) -> tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
index | ::mlir::IntegerAttr | 32-bit signless integer attribute whose value is non-negative |
Operands:
Operand | תֵאוּר |
---|---|
operand | nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
mhlo.if
(mhlo::IfOp)
If operation
Produces the output from executing exactly one branch from true_branch
or false_branch
depending on the value of pred
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#if
Example: %result = "mhlo.if"(%pred) ({ "mhlo.return"(%result_true_branch) : (tensor
Traits: RecursiveMemoryEffects
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
Interfaces: InferTypeOpInterface
Operands:
Operand | תֵאוּר |
---|---|
pred | ranked tensor of pred (AKA boolean or 1-bit integer) values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.imag
(mhlo::ImagOp)
Imag operation
תַחבִּיר:
operation ::= `mhlo.imag` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Extracts the imaginary part, element-wise, from the operand
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#imag
דוּגמָה:
%result = mhlo.imag %operand : (tensor<2xcomplex<f32>>) -> tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.infeed
(mhlo::InfeedOp)
Infeed operation
Reads data from the infeed and produces results
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#infeed
דוּגמָה:
%results:2 = "mhlo.infeed"(%token) {
infeed_config = ""
} : (!mhlo.token) -> (tensor<3x3x3xi32>, !mhlo.token)
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
infeed_config | ::mlir::StringAttr | string attribute |
layout | ::mlir::ArrayAttr | array attribute |
Operands:
Operand | תֵאוּר |
---|---|
token | אֲסִימוֹן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.iota
(mhlo::IotaOp)
Iota operation
Fills an output
tensor with values in increasing order starting from zero along the iota_dimension
dimension.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#iota
דוּגמָה:
%output = mhlo.iota dim = 0 : tensor<4x5xi32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
iota_dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is non-negative |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | statically shaped tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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
(mhlo::IsFiniteOp)
IsFinite operation
תַחבִּיר:
operation ::= `mhlo.is_finite` $x attr-dict `:` functional-type(operands, results)
Performs element-wise check whether the value in x
is finite (ie is neither +Inf, -Inf, nor NaN) and produces a y
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#is_finite
דוּגמָה:
%y = mhlo.is_finite %x : (tensor<7xf32>) -> tensor<7xi1>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
x | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
y | ranked tensor of pred (AKA boolean or 1-bit integer) values |
mhlo.log
(mhlo::LogOp)
Log operation
תַחבִּיר:
operation ::= `mhlo.log` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise logarithm operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#log
דוּגמָה:
%result = mhlo.log %operand : tensor<2x2xf64>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.log_plus_one
(mhlo::Log1pOp)
Log1p operation
תַחבִּיר:
operation ::= `mhlo.log_plus_one` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise logarithm plus one operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#log_plus_one
דוּגמָה:
%result = mhlo.log_plus_one %operand : tensor<6xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.logistic
(mhlo::LogisticOp)
Logistic operation
תַחבִּיר:
operation ::= `mhlo.logistic` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise logistic operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#logistic
דוּגמָה:
%result = mhlo.logistic %operand : tensor<2x2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.map
(mhlo::MapOp)
Map operation
Applies a map function computation
to inputs
along the dimensions
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#map
דוּגמָה:
%result = "mhlo.map"(%input0, %input1) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = mhlo.multiply %arg0, %arg1 : tensor<i32>
mhlo.return %0 : tensor<i32>
}) {
dimensions = dense<[0, 1]> : tensor<2xi64>
} : (tensor<2x2xi32>, tensor<2x2xi32>) -> tensor<2x2xi32>
Traits: InferTensorType
, RecursiveMemoryEffects
, SameOperandsAndResultShape
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimensions | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.maximum
(mhlo::MaxOp)
Max operation
תַחבִּיר:
operation ::= `mhlo.maximum` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise max operation on tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#maximum
דוּגמָה:
%result = mhlo.maximum %lhs, %rhs : tensor<4xf32>
Traits: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.minimum
(mhlo::MinOp)
Min operation
תַחבִּיר:
operation ::= `mhlo.minimum` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise min operation on tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#minimum
דוּגמָה:
%result = mhlo.minimum %lhs, %rhs : tensor<4xf32>
Traits: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.minimum_broadcast_shapes
(mhlo::MinimumBroadcastShapesOp)
Minimizes the rank of two or more shapes to be broadcasted
תַחבִּיר:
operation ::= `mhlo.minimum_broadcast_shapes` $shapes attr-dict `:` type($shapes) `->` type($results)
Given two or more 1D tensors representing shapes, returns one 1D tensor for each operand, where operand i
corresponds to output i
.
The returned tensors have the property that they specify a shape which is a reshape of the corresponding input shape, and the broadcasted output shape (using shape::BroadcastOp) of the returned shapes is a reshape of the broadcasted output shape of the input shapes. Among all possibilities with this property, the one is chosen which minimizes the rank of each returned shape.
The general idea of this op is that it can be used for ops which have a broadcasting semantic to operate on shapes with a possibly smaller rank while preserving equivalence of the computed values. After computing the result of the op using reshaped operands, the result can be reshaped to the result that would have been originally computed.
Here is an example with two input shapes:
mhlo.minimum_broadcast_shapes [1, 2, 3, 1, 2, 1],
[1, 1, 1, 2, 3] -> [6, 2, 1], [2, 3]
The broadcasted output shape of the operands is [1, 2, 3, 1, 2, 3], the broadcasted output shape of the outputs is [6, 2, 3]. These two shapes are reshapes of each other, and also each output is a reshape of the corresponding input.
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
shapes | variadic of 1D tensor of index values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
results | variadic of 1D tensor of index values |
mhlo.multiply
(mhlo::MulOp)
Mul operation
תַחבִּיר:
operation ::= `mhlo.multiply` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise product of two tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#multiply
דוּגמָה:
%result = mhlo.multiply %lhs, %rhs : tensor<2xi32>
Traits: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.negate
(mhlo::NegOp)
Neg operation
תַחבִּיר:
operation ::= `mhlo.negate` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise negation of operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#negate
דוּגמָה:
%result = mhlo.negate %operand : tensor<2x3xi32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.not
(mhlo::NotOp)
Not operation
תַחבִּיר:
operation ::= `mhlo.not` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise NOT of tensor operand
of type integer and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#not
דוּגמָה:
%result = mhlo.not %operand : tensor<5x3x1xi1>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.optimization_barrier
(mhlo::OptimizationBarrierOp)
OptimizationBarrier operation
תַחבִּיר:
operation ::= `mhlo.optimization_barrier` attr-dict ($operand^ `:` custom<PairwiseOpType>(type($operand), type($result))):(`(` `)`)?
Ensures that the operations that produce the operand
are executed before any operations that depend on the result
and prevents compiler transformations from moving operations across the barrier. Other than that, the operation is an identity, ie result
= operand
.
See https://github.com/openxla/stablehlo/blob/main/docs/spec.md#optimization_barrier
דוּגמָה:
%result0, %result1 = mhlo.optimization_barrier %operand0, %operand1 : tensor<f32>, tensor<f32>
Traits: AlwaysSpeculatableImplTrait
, HLO_PairwiseSameOperandAndResultType
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.or
(mhlo::OrOp)
Or operation
תַחבִּיר:
operation ::= `mhlo.or` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise OR of two tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#or
דוּגמָה:
%result = mhlo.or %lhs, %rhs : tensor<2xi1>
Traits: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
rhs | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.outfeed
(mhlo::OutfeedOp)
Outfeed operation
Writes inputs
to the outfeed and produces a result
token.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#outfeed
דוּגמָה:
%result = "mhlo.outfeed"(%input0, %token) {
outfeed_config = ""
} : (tensor<3x3x3xi32>, !mhlo.token) -> !mhlo.token
Interfaces: InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
outfeed_config | ::mlir::StringAttr | string attribute |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
token | אֲסִימוֹן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | אֲסִימוֹן |
mhlo.pad
(mhlo::PadOp)
Pad operation
Expands operand
by padding around the tensor as well as between the elements of the tensor with the given padding_value
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#pad
דוּגמָה:
%0 = mhlo.pad %arg0, %arg1, low = [0, 1], high = [2, 1], interior = [1, 2]
: (tensor<2x3xi32>, tensor<i32>) -> tensor<5x9xi32>
Traits: AlwaysSpeculatableImplTrait
, SameOperandsAndResultElementType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
padding_value | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.partition_id
(mhlo::PartitionIdOp)
PartitionId operation
תַחבִּיר:
operation ::= `mhlo.partition_id` attr-dict `:` type(results)
Produces partition_id
of the current process.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#partition_id
דוּגמָה:
%result = mhlo.partition_id : tensor<ui32>
Interfaces: InferTypeOpInterface
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of 32-bit unsigned integer values |
mhlo.popcnt
(mhlo::PopulationCountOp)
PopulationCount operation
תַחבִּיר:
operation ::= `mhlo.popcnt` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise count of the number of bits set in the operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#popcnt
דוּגמָה:
%result = mhlo.popcnt %operand : tensor<4xi8>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.power
(mhlo::PowOp)
Pow operation
תַחבִּיר:
operation ::= `mhlo.power` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise exponentiation of lhs
tensor by rhs
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#power
דוּגמָה:
%result = mhlo.power %lhs, %rhs : tensor<6xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.real
(mhlo::RealOp)
Real operation
תַחבִּיר:
operation ::= `mhlo.real` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Extracts the real part, element-wise, from the operand
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#real
דוּגמָה:
%result = mhlo.real %operand : (tensor<2xcomplex<f32>>) -> tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.real_dynamic_slice
(mhlo::RealDynamicSliceOp)
RealDynamicSlice operation
תַחבִּיר:
operation ::= `mhlo.real_dynamic_slice` operands attr-dict `:` functional-type(operands, results)
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/8
Informally, this operation does the same thing as SliceOp except that start_indices
, limit_indices
and strides
are specified dynamically: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#slice
דוּגמָה:
%result = mhlo.real_dynamic_slice %operand,
%start_indices, %limit_indices, %strides
: (tensor<256x?xf32>, tensor<2xindex>, tensor<2xindex>, tensor<2xindex>) -> tensor<256x?xf32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
start_indices | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
limit_indices | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
strides | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.recv
(mhlo::RecvOp)
Recv operation
Receives data from a channel with channel_id
and produces results
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#recv
דוּגמָה:
%results:2 = "mhlo.recv"(%token) {
// channel_id = 5 : i64,
// channel_type = #stablehlo<channel_type HOST_TO_DEVICE>,
channel_handle = #mhlo.channel_handle<handle = 5, type = 3>,
is_host_transfer = true
} : (!mhlo.token) -> (tensor<3x4xi32>, !mhlo.token)
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
channel_handle | ::mlir::mhlo::ChannelHandleAttr | two 64-bit integers 'handle' and 'type' |
is_host_transfer | ::mlir::BoolAttr | bool attribute |
Operands:
Operand | תֵאוּר |
---|---|
token | אֲסִימוֹן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.reduce
(mhlo::ReduceOp)
Reduce operation
Applies a reduction function body
to inputs
and init_values
along the dimensions
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reduce
דוּגמָה:
%result = "mhlo.reduce"(%input, %init_value) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.add"(%arg0, %arg1) : (tensor<i32>, tensor<i32>) -> tensor<i32>
"mhlo.return"(%0) : (tensor<i32>) -> ()
}) {
dimensions = dense<1> : tensor<1xi64>
} : (tensor<1x6xi32>, tensor<i32>) -> tensor<1xi32>
Traits: InferTensorType
, RecursiveMemoryEffects
, SameVariadicOperandSize
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimensions | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
init_values | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.reduce_precision
(mhlo::ReducePrecisionOp)
ReducePrecision operation
תַחבִּיר:
operation ::= `mhlo.reduce_precision` $operand `,` `format` `=` custom<ExponentMantissa>($exponent_bits, $mantissa_bits)
attr-dict `:` custom<SameOperandsAndResultType>(type($operand), type($output))
Performs element-wise conversion of operand
to another floating-point type that uses exponent_bits
and mantissa_bits
and back to the original floating-point type and produces an output
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reduce_precision
דוּגמָה:
%output = mhlo.reduce_precision %operand, format = e5m2 : tensor<6xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
exponent_bits | ::mlir::IntegerAttr | 32-bit signless integer attribute whose value is positive |
mantissa_bits | ::mlir::IntegerAttr | 32-bit signless integer attribute whose value is non-negative |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.reduce_scatter
(mhlo::ReduceScatterOp)
ReduceScatter operation
Within each process group in the process grid, performs reduction, using computations
, over the values of the operand
tensor from each process, splits the reduction result along scatter_dimension
into parts, and scatters the split parts between the processes to produce the result
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reduce_scatter
דוּגמָה:
%result = "mhlo.reduce_scatter"(%operand) ({
^bb0(%arg0: tensor<f32>, %arg1: tensor<f32>):
%0 = mhlo.add %arg0, %arg1 : tensor<f32>
mhlo.return %0 : tensor<f32>
}) {
scatter_dimension = 1 : i64,
replica_groups = dense<[[0, 1]]> : tensor<1x2xi64>,
// channel_id = 0
channel_handle = #mhlo.channel_handle<handle = 0, type = 0>
// use_global_device_ids = false
} : (tensor<2x4xf32>) -> tensor<2x2xf32>
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
scatter_dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is non-negative |
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.reduce_window
(mhlo::ReduceWindowOp)
ReduceWindow operation
Applies a reduction function body
to windows of inputs
and init_values
and produces results
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reduce_window
דוּגמָה:
%result = "mhlo.reduce_window"(%input, %init_value) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = mhlo.add %arg0, %arg1 : tensor<i32>
mhlo.return %0 : tensor<i32>
}) {
window_dimensions = dense<[2, 1]> : tensor<2xi64>,
window_strides = dense<[4, 1]> : tensor<2xi64>,
base_dilations = dense<[2, 1]> : tensor<2xi64>,
window_dilations = dense<[3, 1]> : tensor<2xi64>,
padding = dense<[[2, 1], [0, 0]]> : tensor<2x2xi64>
} : (tensor<3x2xi32>, tensor<i32>) -> tensor<2x2xi32>
Traits: InferTensorType
, RecursiveMemoryEffects
, SameVariadicOperandSize
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
init_values | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.remainder
(mhlo::RemOp)
Rem operation
תַחבִּיר:
operation ::= `mhlo.remainder` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise remainder of dividend lhs
and divisor rhs
tensors and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#remainder
דוּגמָה:
%result = mhlo.remainder %lhs, %rhs : tensor<4xi64>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.replica_id
(mhlo::ReplicaIdOp)
ReplicaId operation
תַחבִּיר:
operation ::= `mhlo.replica_id` attr-dict `:` type(results)
Produces replica_id
of the current process.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#replica_id
דוּגמָה:
%result = mhlo.replica_id : tensor<ui32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of 32-bit unsigned integer values |
mhlo.reshape
(mhlo::ReshapeOp)
Reshape operation
תַחבִּיר:
operation ::= `mhlo.reshape` operands attr-dict `:` functional-type(operands, results)
Performs reshape of operand
tensor to a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reshape
דוּגמָה:
%result = mhlo.reshape %operand : (tensor<2xf32>) -> tensor<1x2xf32>
Traits: AlwaysSpeculatableImplTrait
, HLO_CompatibleOperandsAndResultElementType
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | statically shaped tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.return
(mhlo::ReturnOp)
_This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/425
Informally, this operation serves as a terminator for regions defined by
the StableHLO ops. Non-StableHLO ops, e.g. `func.func`, have their own
terminators, e.g. `func.return`.
Example:
```mlir
%result = "mhlo.reduce"(%input, %init_value) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.add"(%arg0, %arg1) : (tensor<i32>, tensor<i32>) -> tensor<i32>
"mhlo.return"(%0) : (tensor<i32>) -> ()
}) {
dimensions = dense<1> : tensor<1xi64>
} : (tensor<1x6xi32>, tensor<i32>) -> tensor<1xi32>
```_
Syntax:
```
operation ::= mhlo.return
$results attr-dict ( :
type($results)^)?
Traits: `AlwaysSpeculatableImplTrait`, `Terminator`
Interfaces: `ConditionallySpeculatable`, `NoMemoryEffect (MemoryEffectOpInterface)`
Effects: `MemoryEffects::Effect{}`
#### Operands:
| Operand | Description |
| :-----: | ----------- |
| `results` | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values
### `mhlo.reverse` (mhlo::ReverseOp)
_Reverse operation_
Reverses the order of elements in the `operand` along the specified
`dimensions` and produces a `result` tensor.
See:
<a href="https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reverse">https://github.com/openxla/stablehlo/blob/main/docs/spec.md#reverse</a>
Example:
```mlir
%result = mhlo.reverse %operand, dims = [1] : tensor<3x2xi32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimensions | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.rng
(mhlo::RngOp)
Rng operation
Generates random numbers using the rng_distribution
algorithm and produces a result
tensor of a given shape shape
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#rng
דוּגמָה:
%result = mhlo.rng %a, %b, %shape, distribution = NORMAL : (tensor<i32>, tensor<i32>, tensor<2xi64>) -> tensor<3x3xi32>
Traits: InferTensorType
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
rng_distribution | ::mlir::mhlo::RngDistributionAttr | XLA PRNG distribution to be used. |
Operands:
Operand | תֵאוּר |
---|---|
a | 0D tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
shape | 1D tensor of index or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.rng_bit_generator
(mhlo::RngBitGeneratorOp)
RngBitGenerator operation
Returns an output
filled with uniform random data and an updated output state output_state
given an initial state initial_state
using the pseudorandom number generator algorithm rng_algorithm
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#rng_bit_generator
דוּגמָה:
%output_state, %output = mhlo.rng_bit_generator %initial_state, algorithm = THREE_FRY : (tensor<2xui64>) -> (tensor<2xui64>, tensor<2x2xui64>)
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
rng_algorithm | ::mlir::mhlo::RngAlgorithmAttr | XLA PRNG algorithm to be used. |
Operands:
Operand | תֵאוּר |
---|---|
initial_state | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
output_state | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
output | statically shaped tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.round_nearest_afz
(mhlo::RoundOp)
Round operation
תַחבִּיר:
operation ::= `mhlo.round_nearest_afz` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise rounding towards the nearest integer, breaking ties away from zero, on the operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#round_nearest_afz
דוּגמָה:
%result = mhlo.round_nearest_afz %operand : tensor<5xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.round_nearest_even
(mhlo::RoundNearestEvenOp)
RoundNearestEven operation
תַחבִּיר:
operation ::= `mhlo.round_nearest_even` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise rounding towards the nearest integer, breaking ties towards the even integer, on the operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#round_nearest_even
דוּגמָה:
%result = mhlo.round_nearest_even %operand : tensor<5xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.rsqrt
(mhlo::RsqrtOp)
Rsqrt operation
תַחבִּיר:
operation ::= `mhlo.rsqrt` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise reciprocal square root operation on operand
tensor and produces a result
tensor, implementing the rSqrt
operation from the IEEE-754 specification.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#rsqrt
דוּגמָה:
%result = mhlo.rsqrt %operand : tensor<2x2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.scatter
(mhlo::ScatterOp)
Scatter operation
Produces results
tensors which are equal to inputs
tensors except that several slices specified by scatter_indices
are updated with the values updates
using update_computation
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#scatter
דוּגמָה:
%result = "mhlo.scatter"(%input, %scatter_indices, %update) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = mhlo.add %arg0, %arg1 : tensor<i32>
mhlo.return %0 : tensor<i32>
}) {
scatter_dimension_numbers = #mhlo.scatter<
update_window_dims = [3, 4],
inserted_window_dims = [1],
input_batching_dims = [0],
scatter_indices_batching_dims = [1],
scatter_dims_to_operand_dims = [2, 1],
index_vector_dim = 3>,
indices_are_sorted = false,
unique_indices = false
} : (tensor<2x3x4x2xi64>, tensor<2x2x3x2xi64>, tensor<2x2x3x2x2xi64>) -> tensor<2x3x4x2xi64>
Traits: RecursiveMemoryEffects
, SameVariadicOperandSize
Interfaces: InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
scatter_indices | ranked tensor of integer or index values |
updates | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.select
(mhlo::SelectOp)
Select operation
תַחבִּיר:
operation ::= `mhlo.select` operands attr-dict `:`
custom<SelectOpType>(type($pred), type($on_true), type($on_false), type($result))
Produces a result
tensor where each element is selected from on_true
or on_false
tensor based on the value of the corresponding element of pred
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#select
דוּגמָה:
%result = mhlo.select %pred, %on_true, %on_false : tensor<2x2xi1>, tensor<2x2xi32>
Traits: AlwaysSpeculatableImplTrait
, HLO_BroadcastingElementwise
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
pred | ranked tensor of pred (AKA boolean or 1-bit integer) values |
on_true | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
on_false | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.select_and_scatter
(mhlo::SelectAndScatterOp)
SelectAndScatter operation
Scatters the values from the source
tensor using scatter
based on the outcome of reduce_window
of the input
tensor using select
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#select_and_scatter
דוּגמָה:
%result = "mhlo.select_and_scatter"(%operand, %source, %init_value) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.compare"(%arg0, %arg1) {
comparison_direction = #stablehlo<comparison_direction GE>
} : (tensor<i32>, tensor<i32>) -> tensor<i1>
"mhlo.return"(%0) : (tensor<i1>) -> ()
}, {
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.add"(%arg0, %arg1) : (tensor<i32>, tensor<i32>) -> tensor<i32>
"mhlo.return"(%0) : (tensor<i32>) -> ()
}) {
window_dimensions = dense<[3, 1]> : tensor<2xi64>,
window_strides = dense<[2, 1]> : tensor<2xi64>,
padding = dense<[[0, 1], [0, 0]]> : tensor<2x2xi64>
} : (tensor<4x2xi32>, tensor<2x2xi32>, tensor<i32>) -> tensor<4x2xi32>
Traits: RecursiveMemoryEffects
Interfaces: InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
source | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
init_value | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.send
(mhlo::SendOp)
Send operation
Sends inputs
to a channel channel_id
and produces a result
token.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#send
דוּגמָה:
%result = "mhlo.send"(%operand, %token) {
// channel_id = 5 : i64,
// channel_type = #stablehlo<channel_type DEVICE_TO_HOST>,
channel_handle = #mhlo.channel_handle<handle = 5, type = 2>,
is_host_transfer = true
} : (tensor<3x4xi32>, !mhlo.token) -> !mhlo.token
Interfaces: InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
channel_handle | ::mlir::mhlo::ChannelHandleAttr | two 64-bit integers 'handle' and 'type' |
is_host_transfer | ::mlir::BoolAttr | bool attribute |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
token | אֲסִימוֹן |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | אֲסִימוֹן |
mhlo.set_dimension_size
(mhlo::SetDimensionSizeOp)
SetDimensionSize operation
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/8
Informally, this operation does the same thing as XLA's SetDimensionSize: https://www.tensorflow.org/xla/operation_semantics#setdimensionsize
דוּגמָה:
%0 = mhlo.set_dimension_size %arg0, %arg1, dim = 1 : (tensor<4x2xf32>, tensor<i32>) -> tensor<4x2xf32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute whose value is non-negative |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
size | tensor of 32-bit signless integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.shift_left
(mhlo::ShiftLeftOp)
ShiftLeft operation
תַחבִּיר:
operation ::= `mhlo.shift_left` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise left-shift operation on the lhs
tensor by rhs
number of bits and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#shift_left
דוּגמָה:
%result = mhlo.shift_left %lhs, %rhs : tensor<6xi8>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.shift_right_arithmetic
(mhlo::ShiftRightArithmeticOp)
ShiftRightArithmetic operation
תַחבִּיר:
operation ::= `mhlo.shift_right_arithmetic` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise arithmetic right-shift operation on the lhs
tensor by rhs
number of bits and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#shift_right_arithmetic
דוּגמָה:
%result = mhlo.shift_right_arithmetic %lhs, %rhs : tensor<6xi8>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.shift_right_logical
(mhlo::ShiftRightLogicalOp)
ShiftRightLogical operation
תַחבִּיר:
operation ::= `mhlo.shift_right_logical` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise logical right-shift operation on the lhs
tensor by rhs
number of bits and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#shift_right_logical
דוּגמָה:
%result = mhlo.shift_right_logical %lhs, %rhs : tensor<6xi8>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
mhlo.sign
(mhlo::SignOp)
Sign operation
תַחבִּיר:
operation ::= `mhlo.sign` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Returns the sign of the operand
element-wise and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#sign
דוּגמָה:
%result = mhlo.sign %operand : tensor<7xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of 2/4/8/16/32/64-bit signless integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.sine
(mhlo::SineOp)
Sine operation
תַחבִּיר:
operation ::= `mhlo.sine` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise sine operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#sine
דוּגמָה:
%result = mhlo.sine %operand : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.slice
(mhlo::SliceOp)
Slice operation
Extracts a slice from the operand
using statically-computed starting indices and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#slice
דוּגמָה:
%result = "mhlo.slice" (%operand) {
start_indices = dense<[1, 2]> : tensor<2xi64>,
limit_indices = dense<[3, 4]> : tensor<2xi64>,
strides = dense<1> : tensor<2xi64>
} : (tensor<3x4xi64>) -> tensor<2x2xi64>
Traits: AlwaysSpeculatableImplTrait
, SameOperandsAndResultElementType
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.sort
(mhlo::SortOp)
Sort operation
Sorts a variadic number of tensors in inputs
together, according to a custom comparator
, along the given dimension
and produces a variadic number of tensors as results
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#sort
דוּגמָה:
%result0, %result1 = "mhlo.sort"(%input0, %input1) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>, %arg2: tensor<i32>, %arg3: tensor<i32>):
%predicate = "mhlo.compare"(%arg0, %arg1) {
comparison_direction = #stablehlo<comparison_direction GT>
} : (tensor<i32>, tensor<i32>) -> tensor<i1>
"mhlo.return"(%predicate) : (tensor<i1>) -> ()
}) {
dimension = 0 : i64,
is_stable = true
} : (tensor<2x3xi32>, tensor<2x3xi32>) -> (tensor<2x3xi32>, tensor<2x3xi32>)
Traits: InferTensorType
, RecursiveMemoryEffects
, SameOperandsAndResultShape
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dimension | ::mlir::IntegerAttr | 64-bit signless integer attribute |
is_stable | ::mlir::BoolAttr | bool attribute |
Operands:
Operand | תֵאוּר |
---|---|
inputs | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.sparse_dot
(mhlo::SparseDotOp)
Sparse dot operation
Similar to dot_general
operation, with one or both of the operands being sparse. An additional argument provides sparsity meta information. Disclaimer: this op is experimental / a work in progress.
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
lhs_sparsity | ::mlir::mhlo::SparsityDescriptorAttr | Describes structured (N:M) sparsity configuration |
rhs_sparsity | ::mlir::mhlo::SparsityDescriptorAttr | Describes structured (N:M) sparsity configuration |
dot_dimension_numbers | ::mlir::mhlo::DotDimensionNumbersAttr | Attribute that models the dimension information for dot. |
precision_config | ::mlir::ArrayAttr | Precision Config attribute |
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
meta | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.sqrt
(mhlo::SqrtOp)
Sqrt operation
תַחבִּיר:
operation ::= `mhlo.sqrt` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise square root operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#sqrt
דוּגמָה:
%result = mhlo.sqrt %operand : tensor<2x2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.stochastic_convert
(mhlo::StochasticConvertOp)
StochasticConvert operation
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/295
Informally, this operation performs element-wise conversion of values from a bigger type to a smaller one with stochastic rounding using the random number passed in.
Traits: AlwaysSpeculatableImplTrait
, Elementwise
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
random | ranked tensor of 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.subtract
(mhlo::SubtractOp)
Subtract operation
תַחבִּיר:
operation ::= `mhlo.subtract` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise subtraction of two tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#subtract
דוּגמָה:
%result = mhlo.subtract %lhs, %rhs : tensor<2xi32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
rhs | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.tan
(mhlo::TanOp)
Tan operation
תַחבִּיר:
operation ::= `mhlo.tan` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
This operation is a work in progress, so it is not yet included in the specification: https://github.com/openxla/stablehlo/issues/954
Informally, this operation returns Tan(operand)
element-wise.
דוּגמָה:
%0 = mhlo.tan %arg0 : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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.tanh
(mhlo::TanhOp)
Tanh operation
תַחבִּיר:
operation ::= `mhlo.tanh` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise hyperbolic tangent operation on operand
tensor and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#tanh
דוּגמָה:
%result = mhlo.tanh %operand : tensor<2xf32>
Traits: AlwaysSpeculatableImplTrait
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values |
mhlo.topk
(mhlo::TopKOp)
TopK operation
תַחבִּיר:
operation ::= `mhlo.topk` `(`$operand `,` `k` `=` $k (`,` `largest` `=` $largest^)? `)` attr-dict `:`
type($operand) `->` `(`type($values)`,` type($indices)`)`
Returns top k
values and their indices, along the last dimension of the operand if largest=true
or the bottom k
values if largest=false
.
See: https://www.tensorflow.org/xla/operation_semantics#top-k
דוּגמָה:
%values, %indices = mhlo.topk(%operand, k=5, largest=true)
: tensor<100xf32> -> (tensor<5xf32>, tensor<5xi32>)
Traits: InferTensorType
, RecursiveMemoryEffects
Interfaces: InferShapedTypeOpInterface
, InferTypeOpInterface
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
k | ::mlir::IntegerAttr | 64-bit signless integer attribute |
largest | ::mlir::BoolAttr | bool attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
values | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
indices | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.torch_index_select
(mhlo::TorchIndexSelectOp)
TorchIndexSelect operation
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/3
Informally, this operation does the same thing as PyTorch's index_select, augmented with support for batch dimensions: https://pytorch.org/docs/stable/generated/torch.index_select.html
The batch_dims
attribute specifies the number of major batch dimensions (0 or more) that act like a multidimensional loop over both the operand and the index.
דוּגמָה:
%result = "mhlo.torch_index_select"(%operand, %index) {
dim = 2 : i64,
batch_dims = 1 : i64
} : (tensor<8x128x3072x64xf32>, tensor<8x16x1024xi32>) -> tensor<8x128x16x1024x64xf32>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
dim | ::mlir::IntegerAttr | 64-bit signless integer attribute |
batch_dims | ::mlir::IntegerAttr | 64-bit signless integer attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
index | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.trace
(mhlo::TraceOp)
Trace operation
תַחבִּיר:
operation ::= `mhlo.trace` $operand `,` $tag attr-dict `:` type($operand)
This operation is on its way out of StableHLO, so it is not included in the specification: https://github.com/openxla/stablehlo/issues/604
It is not used by JAX, PyTorch or TensorFlow, so it looks like we should've classified it as "Private to XLA" and not included it in StableHLO in the first place. With that in mind, its semantics will not be documented here.
דוּגמָה:
mhlo.trace %arg0, "In test code." : tensor<5x1x5xi32>
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
tag | ::mlir::StringAttr | string attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.transpose
(mhlo::TransposeOp)
Transpose operation
Permutes the dimensions of operand
tensor using permutation
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#transpose
דוּגמָה:
%0 = mhlo.transpose %arg0, dims = [2, 1, 0] : (tensor<1x2x3xi32>) -> tensor<3x2x1xi32>
Traits: AlwaysSpeculatableImplTrait
, HLO_CompatibleOperandsAndResultElementType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
permutation | ::mlir::DenseIntElementsAttr | 64-bit signless integer elements attribute |
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.triangular_solve
(mhlo::TriangularSolveOp)
TriangularSolve operation
Solves batches of systems of linear equations with lower or upper triangular coefficient matrices.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#triangular_solve
דוּגמָה:
%result = "mhlo.triangular_solve"(%a, %b) {
left_side = true,
lower = true,
unit_diagonal = false,
transpose_a = #stablehlo<transpose NO_TRANSPOSE>
} : (tensor<3x3xf32>, tensor<3x3xf32>) -> tensor<3x3xf32>
Traits: AlwaysSpeculatableImplTrait
, InferTensorType
, SameOperandsAndResultElementType
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Attributes:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
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 | תֵאוּר |
---|---|
a | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 |
b | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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 |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type 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.tuple
(mhlo::TupleOp)
Tuple operation
תַחבִּיר:
operation ::= `mhlo.tuple` $val attr-dict `:` custom<TupleOpType>(type($val), type($result))
Produces a result
tuple from values val
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#tuple
דוּגמָה:
%result = mhlo.tuple %val0, %val1 : tuple<tensor<2xf32>, tuple<tensor<i32>>>
Traits: AlwaysSpeculatableImplTrait
Interfaces: ConditionallySpeculatable
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
val | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token or nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | nested tuple with any combination of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token values |
mhlo.uniform_dequantize
(mhlo::UniformDequantizeOp)
UniformDequantize operation
תַחבִּיר:
operation ::= `mhlo.uniform_dequantize` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise conversion of quantized tensor operand
to a floating-point tensor result
according to the quantization parameters defined by the operand
type.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#uniform_dequantize
דוּגמָה:
%result = mhlo.uniform_dequantize %operand : (tensor<16x16x!quant.uniform<i8:f32, 34.0:16>>) -> tensor<16x16xf32>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, InferTensorType
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type values |
mhlo.uniform_quantize
(mhlo::UniformQuantizeOp)
UniformQuantize operation
תַחבִּיר:
operation ::= `mhlo.uniform_quantize` $operand attr-dict
`:` custom<SameOperandsAndResultType>(type($operand), type($result))
Performs element-wise conversion of floating-point tensor or quantized tensor operand
to a quantized tensor result
according to the quantization parameters defined by the result
type.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#uniform_quantize
דוּגמָה:
%result = mhlo.uniform_quantize %operand : (tensor<16x16xf32>) -> tensor<16x16x!quant.uniform<ui8:f32, 34.0:16>>
Traits: AlwaysSpeculatableImplTrait
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
operand | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
mhlo.while
(mhlo::WhileOp)
While operation
Produces the output from executing body
function 0 or more times while the cond
function outputs true
.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#while
דוּגמָה:
%results0, %results1 = "mhlo.while"(%operand0, %operand1) ({
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.compare"(%arg0, %arg1) {
comparison_direction = #stablehlo<comparison_direction LT>
} : (tensor<i32>, tensor<i32>) -> tensor<i1>
"mhlo.return"(%0) : (tensor<i1>) -> ()
}, {
^bb0(%arg0: tensor<i32>, %arg1: tensor<i32>):
%0 = "mhlo.add"(%arg0, %constant0) : (tensor<i32>, tensor<i32>) -> tensor<i32>
"mhlo.return"(%0, %arg1) : (tensor<i32>, tensor<i32>) -> ()
}) : (tensor<i32>, tensor<i32>) -> (tensor<i32>, tensor<i32>)
Traits: RecursiveMemoryEffects
, SingleBlockImplicitTerminator<ReturnOp>
, SingleBlock
Interfaces: InferTypeOpInterface
, OpAsmOpInterface
Operands:
Operand | תֵאוּר |
---|---|
operand | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | variadic of ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer values or ranked tensor of 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values or token |
mhlo.xla.rng_get_and_update_state
(mhlo::XlaRngGetAndUpdateStateOp)
XlaRngGetAndUpdateState operation
תַחבִּיר:
operation ::= `mhlo.xla.rng_get_and_update_state` attr-dict
This operation is private to the XLA compiler, so it is does not yet have a specification.
Informally, this operation 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:
תְכוּנָה | MLIR Type | תֵאוּר |
---|---|---|
delta | ::mlir::IntegerAttr | 64-bit signless integer attribute |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
«unnamed» | statically shaped tensor of 64-bit unsigned integer values |
mhlo.xor
(mhlo::XorOp)
Xor operation
תַחבִּיר:
operation ::= `mhlo.xor` $lhs `,` $rhs attr-dict
`:` custom<SameOperandsAndResultType>(type($lhs), type($rhs), type($result))
Performs element-wise XOR of two tensors lhs
and rhs
and produces a result
tensor.
See: https://github.com/openxla/stablehlo/blob/main/docs/spec.md#xor
דוּגמָה:
%result = mhlo.xor %lhs, %rhs : tensor<2xi32>
Traits: AlwaysSpeculatableImplTrait
, Commutative
, CompatibleOperandsAndResultType
, Elementwise
, SameOperandsAndResultShape
Interfaces: ConditionallySpeculatable
, InferShapedTypeOpInterface
, InferTypeOpInterface
, NoMemoryEffect (MemoryEffectOpInterface)
Effects: MemoryEffects::Effect{}
Operands:
Operand | תֵאוּר |
---|---|
lhs | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
rhs | ranked tensor of pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer values |
תוצאות:
תוֹצָאָה | תֵאוּר |
---|---|
result | ranked tensor of f4E2M1FN type or f6E2M3FN type or f6E3M2FN type or f8E3M4 type or f8E4M3 type or f8E4M3FN type or f8E4M3FNUZ type or f8E4M3B11FNUZ type or f8E5M2 type or f8E5M2FNUZ type or f8E8M0FNU type or 16-bit float or 32-bit float or 64-bit float or bfloat16 type or pred (AKA boolean or 1-bit integer) or 2/4/8/16/32/64-bit signless integer or 2/4/8/16/32/64-bit unsigned integer or complex type with 32-bit float or 64-bit float elements or 2/4/8/16/32-bit uniform quantized signed integer or 2/4/8/16/32-bit uniform quantized unsigned integer or 2/4/8/16/32-bit uniform quantized per axis signed integer or 2/4/8/16/32-bit uniform quantized per axis unsigned integer values |
Attributes
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:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
argTupleIndices | ::llvm::ArrayRef<int64_t> | מֵמַד |
resultIndex | int64_t | |
resultTupleIndices | ::llvm::ArrayRef<int64_t> | מֵמַד |
isMustAlias | bool |
ChannelHandleAttr
two 64-bit integers 'handle' and 'type'
תַחבִּיר:
#mhlo.channel_handle<
int64_t, # handle
int64_t # type
>
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
יָדִית | int64_t | |
סוּג | int64_t |
ComparisonDirectionAttr
Which comparison operation to perform.
תַחבִּיר:
#mhlo.comparison_direction<
::mlir::mhlo::ComparisonDirection # value
>
Enum cases:
- EQ (
EQ
) - NE (
NE
) - GE (
GE
) - GT (
GT
) - LE (
LE
) - LT (
LT
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::ComparisonDirection | an enum of type ComparisonDirection |
ComparisonTypeAttr
Which comparison type to use.
תַחבִּיר:
#mhlo.comparison_type<
::mlir::mhlo::ComparisonType # value
>
Enum cases:
- NOTYPE (
NOTYPE
) - FLOAT (
FLOAT
) - TOTALORDER (
TOTALORDER
) - SIGNED (
SIGNED
) - UNSIGNED (
UNSIGNED
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::ComparisonType | an enum of type ComparisonType |
ConvDimensionNumbersAttr
Structure of dimension information for conv op
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
inputBatchDimension | int64_t | |
inputFeatureDimension | int64_t | |
inputSpatialDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
kernelInputFeatureDimension | int64_t | |
kernelOutputFeatureDimension | int64_t | |
kernelSpatialDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
outputBatchDimension | int64_t | |
outputFeatureDimension | int64_t | |
outputSpatialDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
CrossProgramPrefetchAttr
Argument that is prefetched from another program
תַחבִּיר:
#mhlo.cross_program_prefetch<
int64_t, # parameter
::llvm::ArrayRef<int64_t>, # indices
std::optional<int64_t> # offset
>
This attribute captures an argument that is prefetched from another program. For a given CrossProgramPrefetchAttr
, parameter
tells us which argument of the main
function of the module is prefetched, and indices
is a shape index telling us what subshape of that argument is prefetched.
A shape has a subshape iff it is a tuple. In that case, the subshape of the tuple by indices
is the shape achieved after indexing by each element of indices
in turn. For example, the [1,0] subshape of tuple<tuple<token, token>, tuple<tensor<i32>, token>>
is tensor<i32>
.
An empty value for indices
means the whole shape is prefetched.
לְדוּגמָה,
module attributes { mhlo.cross_program_prefetch = [ #mhlo.cross_program_prefetch< parameter = 0, indices = [0]> ]} {
func.func @copy(%arg0 : tuple<tensor<2x3xi32>, tensor<i32>>) -> tuple<tensor<2x3xi32>, tensor<i32>> {
%0 = "mhlo.copy"(%arg0) {is_cross_program_prefetch}
return %0 : tuple<tensor<2x3xi32>, tensor<i32>>
}
func.func @main(%arg0 : tuple<tensor<2x3xi32>, tensor<i32>>) -> tuple<tensor<2x3xi32>, tensor<i32>> {
%1 = "mhlo.async_start"(%arg0) {called_computation=@copy}
%2 = "mhlo.async_done"(%1) {called_computation=@copy}
return %2 : tuple<tensor<2x3xi32>, tensor<i32>>
}
}
The parameter = 0
tells us that the async copy of the 0
th parameter is a cross_program_prefetch
, while the index
of [0]
tells us that the 0
th element of the tuple is prefetched while the other element of the tuple is not.
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
פָּרָמֶטֶר | int64_t | |
indices | ::llvm::ArrayRef<int64_t> | מֵמַד |
לְקַזֵז | std::optional<int64_t> |
CustomCallScheduleAttr
Specifies the desired schedule for the custom-call.
תַחבִּיר:
#mhlo.custom_call_schedule<
::mlir::mhlo::CustomCallSchedule # value
>
Enum cases:
- NONE (
NONE
) - LATEST (
LATEST
) - EARLIEST (
EARLIEST
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::CustomCallSchedule | an enum of type CustomCallSchedule |
DequantizeModeAttr
Dequantization mode. Only MIN_COMBINED is supported.
תַחבִּיר:
#mhlo.dequantize_mode<
::mlir::mhlo::DequantizeMode # value
>
Enum cases:
- MIN_COMBINED (
MIN_COMBINED
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::DequantizeMode | an enum of type DequantizeMode |
DomainKindAttr
Kind of domain metatdata attached to an HLO domain.
תַחבִּיר:
#mhlo.kind<
::mlir::mhlo::DomainKind # value
>
Enum cases:
- sharding (
sharding
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::DomainKind | an enum of type DomainKind |
DotAlgorithmAttr
Attribute that models the algorithm constraints to use for computing dot.
תַחבִּיר:
#mhlo.dot_algorithm<
Type, # lhsPrecisionType
Type, # rhsPrecisionType
Type, # accumulationType
int64_t, # lhsComponentCount
int64_t, # rhsComponentCount
int64_t, # numPrimitiveOperations
bool # allowImpreciseAccumulation
>
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
lhsPrecisionType | Type | |
rhsPrecisionType | Type | |
accumulationType | Type | |
lhsComponentCount | int64_t | |
rhsComponentCount | int64_t | |
numPrimitiveOperations | int64_t | |
allowImpreciseAccumulation | bool |
DotDimensionNumbersAttr
Attribute that models the dimension information for dot.
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
lhsBatchingDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
rhsBatchingDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
lhsContractingDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
rhsContractingDimensions | ::llvm::ArrayRef<int64_t> | מֵמַד |
FftTypeAttr
XLA fast fourier transform type.
תַחבִּיר:
#mhlo.fft_type<
::mlir::mhlo::FftType # value
>
Enum cases:
- FFT (
FFT
) - IFFT (
IFFT
) - RFFT (
RFFT
) - IRFFT (
IRFFT
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::FftType | an enum of type FftType |
FusionKindAttr
fusion kind
תַחבִּיר:
#mhlo.fusion_kind<
::mlir::mhlo::FusionKind # value
>
Enum cases:
- kLoop (
kLoop
) - kInput (
kInput
) - kOutput (
kOutput
) - kCustom (
kCustom
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::FusionKind | an enum of type FusionKind |
GatherDimensionNumbersAttr
Attribute that models the dimension information for gather
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
offsetDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
collapsedSliceDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
operandBatchingDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
startIndicesBatchingDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
startIndexMap | ::llvm::ArrayRef<int64_t> | מֵמַד |
indexVectorDim | int64_t |
OutputOperandAliasAttr
Attribute that models the alias relationship of output and operand of a CustomCall op
תַחבִּיר:
#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:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
outputTupleIndices | ::llvm::ArrayRef<int64_t> | מֵמַד |
operandIndex | int64_t | |
operandTupleIndices | ::llvm::ArrayRef<int64_t> | מֵמַד |
PrecisionAttr
XLA precision for an operand. Has backend specific meaning.
תַחבִּיר:
#mhlo.precision<
::mlir::mhlo::Precision # value
>
Enum cases:
- DEFAULT (
DEFAULT
) - HIGH (
HIGH
) - HIGHEST (
HIGHEST
) - PACKED_NIBBLE (
PACKED_NIBBLE
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::Precision | an enum of type Precision |
RngAlgorithmAttr
XLA PRNG algorithm to be used.
תַחבִּיר:
#mhlo.rng_algorithm<
::mlir::mhlo::RngAlgorithm # value
>
Enum cases:
- DEFAULT (
DEFAULT
) - THREE_FRY (
THREE_FRY
) - PHILOX (
PHILOX
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::RngAlgorithm | an enum of type RngAlgorithm |
RngDistributionAttr
XLA PRNG distribution to be used.
תַחבִּיר:
#mhlo.rng_distribution<
::mlir::mhlo::RngDistribution # value
>
Enum cases:
- UNIFORM (
UNIFORM
) - NORMAL (
NORMAL
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::RngDistribution | an enum of type RngDistribution |
ScatterDimensionNumbersAttr
Attribute that models the dimension information for scatter
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
updateWindowDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
insertedWindowDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
inputBatchingDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
scatterIndicesBatchingDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
scatterDimsToOperandDims | ::llvm::ArrayRef<int64_t> | מֵמַד |
indexVectorDim | int64_t |
SparsityDescriptorAttr
Describes structured (N:M) sparsity configuration
תַחבִּיר:
#mhlo.sparsity<
int64_t, # dimension
int64_t, # n
int64_t # m
>
This attribute is defined for a sparse dot operation with a structured sparse input tensor. With (N=2,M=4), every 4 consecutive logical elements have exactly 2 non-zero physical elements in the input tensor.
$dimension defines the index of the contracting dimension that is sparse (it has to be the most minor dimension). The additional metadata operand in the sparse dot operation defines which logical elements are zeroed out.
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
מֵמַד | int64_t | |
נ | int64_t | |
מ | int64_t |
TransposeAttr
Transpose options
תַחבִּיר:
#mhlo.transpose<
::mlir::mhlo::Transpose # value
>
Enum cases:
- TRANSPOSE_INVALID (
TRANSPOSE_INVALID
) - NO_TRANSPOSE (
NO_TRANSPOSE
) - TRANSPOSE (
TRANSPOSE
) - ADJOINT (
ADJOINT
)
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
עֵרֶך | ::mlir::mhlo::Transpose | an enum of type Transpose |
TypeExtensionsAttr
Attribute that extends tensor type with MHLO type properties.
תַחבִּיר:
#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:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
bounds | ::llvm::ArrayRef<int64_t> |
סוגים
AsyncBundleType
Opaque collection of other types
תַחבִּיר:
!mhlo.async_bundle<
::llvm::ArrayRef<Type> # types
>
Parameters:
פָּרָמֶטֶר | C++ type | תֵאוּר |
---|---|---|
סוגים | ::llvm::ArrayRef<Type> |
Enums
ComparisonDirection
Which comparison operation to perform.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
EQ | 0 | EQ |
NE | 1 | NE |
GE | 2 | GE |
GT | 3 | GT |
LE | 4 | LE |
LT | 5 | LT |
ComparisonType
Which comparison type to use.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
NOTYPE | 0 | NOTYPE |
לָצוּף | 1 | לָצוּף |
TOTALORDER | 2 | TOTALORDER |
חָתוּם | 3 | חָתוּם |
UNSIGNED | 4 | UNSIGNED |
CustomCallApiVersion
Custom call API version
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
API_VERSION_UNSPECIFIED | 0 | API_VERSION_UNSPECIFIED |
API_VERSION_ORIGINAL | 1 | API_VERSION_ORIGINAL |
API_VERSION_STATUS_RETURNING | 2 | API_VERSION_STATUS_RETURNING |
API_VERSION_STATUS_RETURNING_UNIFIED | 3 | API_VERSION_STATUS_RETURNING_UNIFIED |
API_VERSION_TYPED_FFI | 4 | API_VERSION_TYPED_FFI |
CustomCallSchedule
Specifies the desired schedule for the custom-call.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
אַף לֹא אֶחָד | 0 | אַף לֹא אֶחָד |
האחרון | 1 | האחרון |
EARLIEST | 2 | EARLIEST |
DequantizeMode
Dequantization mode. Only MIN_COMBINED is supported.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
MIN_COMBINED | 0 | MIN_COMBINED |
DomainKind
Kind of domain metatdata attached to an HLO domain.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
sharding | 0 | sharding |
FftType
XLA fast fourier transform type.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
FFT | 0 | FFT |
IFFT | 1 | IFFT |
RFFT | 2 | RFFT |
IRFFT | 3 | IRFFT |
FusionKind
fusion kind
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
kLoop | 0 | kLoop |
kInput | 1 | kInput |
kOutput | 2 | kOutput |
kCustom | 3 | kCustom |
דִיוּק
XLA precision for an operand. Has backend specific meaning.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
בְּרִירַת מֶחדָל | 0 | בְּרִירַת מֶחדָל |
גָבוֹהַ | 1 | גָבוֹהַ |
הֲכִי גָבוֹהַ | 2 | הֲכִי גָבוֹהַ |
PACKED_NIBBLE | 3 | PACKED_NIBBLE |
RngAlgorithm
XLA PRNG algorithm to be used.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
בְּרִירַת מֶחדָל | 0 | בְּרִירַת מֶחדָל |
THREE_FRY | 1 | THREE_FRY |
PHILOX | 2 | PHILOX |
RngDistribution
XLA PRNG distribution to be used.
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
אָחִיד | 1 | אָחִיד |
נוֹרמָלִי | 2 | נוֹרמָלִי |
לְשַׁרבֵּב
Transpose options
Cases:
סֵמֶל | עֵרֶך | חוּט |
---|---|---|
TRANSPOSE_INVALID | 0 | TRANSPOSE_INVALID |
NO_TRANSPOSE | 1 | NO_TRANSPOSE |
לְשַׁרבֵּב | 2 | לְשַׁרבֵּב |
ADJOINT | 3 | ADJOINT |