Vector Packed Arithmetic Instructions
279 AMDGPU instructions in this category - showing 100 per page, page 1 of 3 - click any row for encoding, pseudocode, and full documentation.
| Mnemonic | Syntax | Format | Summary |
|---|---|---|---|
| v_dot2_f32_bf16 | v_dot2_f32_bf16 | VOP3P | Calculate the dot product of BF16 float 2-vectors from the first and second inputs, convert the product to single-precision float format, add the… |
| v_dot2_f32_f16 | v_dot2_f32_f16 | VOP3P | Compute the dot product of two packed 2-D half-precision float inputs in the single-precision float domain, add a single-precision float value from… |
| v_dot2_i32_i16 | v_dot2_i32_i16 | VOP3P | Compute the dot product of two packed 2-D signed 16-bit integer inputs in the signed 32-bit integer domain, add a signed 32-bit integer value from… |
| v_dot2_u32_u16 | v_dot2_u32_u16 | VOP3P | Compute the dot product of two packed 2-D unsigned 16-bit integer inputs in the unsigned 32-bit integer domain, add an unsigned 32-bit integer value… |
| v_dot4_f32_bf8_bf8 | v_dot4_f32_bf8_bf8 | VOP3P | Compute the dot product of two packed 4-D BF8 float inputs in the single-precision float domain, add a single-precision float value from the third… |
| v_dot4_f32_bf8_fp8 | v_dot4_f32_bf8_fp8 | VOP3P | Compute the dot product of a packed 4-D BF8 float input and a packed 4-D FP8 float input in the single-precision float domain, add a single-precision… |
| v_dot4_f32_fp8_bf8 | v_dot4_f32_fp8_bf8 | VOP3P | Compute the dot product of a packed 4-D FP8 float input and a packed 4-D BF8 float input in the single-precision float domain, add a single-precision… |
| v_dot4_f32_fp8_fp8 | v_dot4_f32_fp8_fp8 | VOP3P | Compute the dot product of two packed 4-D FP8 float inputs in the single-precision float domain, add a single-precision float value from the third… |
| v_dot4_i32_i8 | v_dot4_i32_i8 | VOP3P | Compute the dot product of two packed 4-D signed 8-bit integer inputs in the signed 32-bit integer domain, add a signed 32-bit integer value from the… |
| v_dot4_i32_iu8 | v_dot4_i32_iu8 | VOP3P | Compute the dot product of two packed 4-D signed or unsigned 8-bit integer inputs in the signed 32-bit integer domain, add a signed 32-bit integer… |
| v_dot4_u32_u8 | v_dot4_u32_u8 | VOP3P | Compute the dot product of two packed 4-D unsigned 8-bit integer inputs in the unsigned 32-bit integer domain, add an unsigned 32-bit integer value… |
| v_dot8_i32_i4 | v_dot8_i32_i4 | VOP3P | Compute the dot product of two packed 8-D signed 4-bit integer inputs in the signed 32-bit integer domain, add a signed 32-bit integer value from the… |
| v_dot8_i32_iu4 | v_dot8_i32_iu4 | VOP3P | Compute the dot product of two packed 8-D signed or unsigned 4-bit integer inputs in the signed 32-bit integer domain, add a signed 32-bit integer… |
| v_dot8_u32_u4 | v_dot8_u32_u4 | VOP3P | Compute the dot product of two packed 8-D unsigned 4-bit integer inputs in the unsigned 32-bit integer domain, add an unsigned 32-bit integer value… |
| v_fma_mix_bf16_t16 | v_fma_mix_bf16_t16 | VOP3P | AMDGPU VOP3P vector instruction. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_fma_mix_f16_t16 | v_fma_mix_f16_t16 | VOP3P | AMDGPU VOP3P vector instruction operating on f16 data. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_fma_mix_f32 | v_fma_mix_f32 | VOP3P | Multiply two inputs and add a third input using fused multiply add where the inputs are a mix of half-precision float and single-precision float… |
| v_fma_mix_f32_bf16 | v_fma_mix_f32_bf16 | VOP3P | AMDGPU VOP3P vector instruction operating on f32 data. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_fma_mixhi_bf16 | v_fma_mixhi_bf16 | VOP3P | AMDGPU VOP3P vector instruction. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_fma_mixhi_f16 | v_fma_mixhi_f16 | VOP3P | Multiply two inputs and add a third input using fused multiply add where the inputs are a mix of half-precision float and single-precision float… |
| v_fma_mixlo_bf16 | v_fma_mixlo_bf16 | VOP3P | AMDGPU VOP3P vector instruction. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_fma_mixlo_f16 | v_fma_mixlo_f16 | VOP3P | Multiply two inputs and add a third input using fused multiply add where the inputs are a mix of half-precision float and single-precision float… |
| v_mad_mix_f32 | v_mad_mix_f32 | VOP3P | Multiply two inputs and add a third input where the inputs are a mix of half-precision float and single- precision float values. |
| v_mad_mixhi_f16 | v_mad_mixhi_f16 | VOP3P | Multiply two inputs and add a third input where the inputs are a mix of half-precision float and single- precision float values. |
| v_mad_mixlo_f16 | v_mad_mixlo_f16 | VOP3P | Multiply two inputs and add a third input where the inputs are a mix of half-precision float and single- precision float values. |
| v_mfma_f32_16x16x128_f8f6f4 | v_mfma_f32_16x16x128_f8f6f4 | VOP3P | Multiply the 16x128 matrix in the first input by the 128x16 matrix in the second input and add the 16x16 matrix in the third input using fused… |
| v_mfma_f32_16x16x16_bf16 | v_mfma_f32_16x16x16_bf16 | VOP3P | Multiply the 16x16 matrix in the first input by the 16x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x16_f16 | v_mfma_f32_16x16x16_f16 | VOP3P | Multiply the 16x16 matrix in the first input by the 16x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x16bf16_1k | v_mfma_f32_16x16x16bf16_1k | VOP3P | Multiply the 16x16 matrix in the first input by the 16x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x1_4b_f32 | v_mfma_f32_16x16x1_4b_f32 | VOP3P | Multiply the 16x1 matrix in the first input by the 1x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x1f32 | v_mfma_f32_16x16x1f32 | VOP3P | Multiply the 16x1 matrix in the first input by the 1x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x2bf16 | v_mfma_f32_16x16x2bf16 | VOP3P | Multiply the 16x2 matrix in the first input by the 2x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_bf16 | v_mfma_f32_16x16x32_bf16 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_bf8_bf8 | v_mfma_f32_16x16x32_bf8_bf8 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_bf8_fp8 | v_mfma_f32_16x16x32_bf8_fp8 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_f16 | v_mfma_f32_16x16x32_f16 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_fp8_bf8 | v_mfma_f32_16x16x32_fp8_bf8 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x32_fp8_fp8 | v_mfma_f32_16x16x32_fp8_fp8 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4_4b_bf16 | v_mfma_f32_16x16x4_4b_bf16 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4_4b_f16 | v_mfma_f32_16x16x4_4b_f16 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4_f32 | v_mfma_f32_16x16x4_f32 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4bf16_1k | v_mfma_f32_16x16x4bf16_1k | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4f16 | v_mfma_f32_16x16x4f16 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x4f32 | v_mfma_f32_16x16x4f32 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x8_xf32 | v_mfma_f32_16x16x8_xf32 | VOP3P | Multiply the 16x8 matrix in the first input by the 8x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_16x16x8bf16 | v_mfma_f32_16x16x8bf16 | VOP3P | Multiply the 16x8 matrix in the first input by the 8x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_bf16 | v_mfma_f32_32x32x16_bf16 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_bf8_bf8 | v_mfma_f32_32x32x16_bf8_bf8 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_bf8_fp8 | v_mfma_f32_32x32x16_bf8_fp8 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_f16 | v_mfma_f32_32x32x16_f16 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_fp8_bf8 | v_mfma_f32_32x32x16_fp8_bf8 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x16_fp8_fp8 | v_mfma_f32_32x32x16_fp8_fp8 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x1_2b_f32 | v_mfma_f32_32x32x1_2b_f32 | VOP3P | Multiply the 32x1 matrix in the first input by the 1x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x1f32 | v_mfma_f32_32x32x1f32 | VOP3P | Multiply the 32x1 matrix in the first input by the 1x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x2_f32 | v_mfma_f32_32x32x2_f32 | VOP3P | Multiply the 32x2 matrix in the first input by the 2x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x2bf16 | v_mfma_f32_32x32x2bf16 | VOP3P | Multiply the 32x2 matrix in the first input by the 2x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x2f32 | v_mfma_f32_32x32x2f32 | VOP3P | Multiply the 32x2 matrix in the first input by the 2x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4_2b_bf16 | v_mfma_f32_32x32x4_2b_bf16 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4_2b_f16 | v_mfma_f32_32x32x4_2b_f16 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4_xf32 | v_mfma_f32_32x32x4_xf32 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4bf16 | v_mfma_f32_32x32x4bf16 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4bf16_1k | v_mfma_f32_32x32x4bf16_1k | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x4f16 | v_mfma_f32_32x32x4f16 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x64_f8f6f4 | v_mfma_f32_32x32x64_f8f6f4 | VOP3P | Multiply the 32x64 matrix in the first input by the 64x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x8_bf16 | v_mfma_f32_32x32x8_bf16 | VOP3P | Multiply the 32x8 matrix in the first input by the 8x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x8_f16 | v_mfma_f32_32x32x8_f16 | VOP3P | Multiply the 32x8 matrix in the first input by the 8x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x8bf16_1k | v_mfma_f32_32x32x8bf16_1k | VOP3P | Multiply the 32x8 matrix in the first input by the 8x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_32x32x8f16 | v_mfma_f32_32x32x8f16 | VOP3P | Multiply the 32x8 matrix in the first input by the 8x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_f32_4x4x1_16b_f32 | v_mfma_f32_4x4x1_16b_f32 | VOP3P | Multiply the 4x1 matrix in the first input by the 1x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x1f32 | v_mfma_f32_4x4x1f32 | VOP3P | Multiply the 4x1 matrix in the first input by the 1x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x2bf16 | v_mfma_f32_4x4x2bf16 | VOP3P | Multiply the 4x2 matrix in the first input by the 2x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x4_16b_bf16 | v_mfma_f32_4x4x4_16b_bf16 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x4_16b_f16 | v_mfma_f32_4x4x4_16b_f16 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x4bf16_1k | v_mfma_f32_4x4x4bf16_1k | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f32_4x4x4f16 | v_mfma_f32_4x4x4f16 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f64_16x16x4_f64 | v_mfma_f64_16x16x4_f64 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f64_16x16x4f64 | v_mfma_f64_16x16x4f64 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_f64_4x4x4_4b_f64 | v_mfma_f64_4x4x4_4b_f64 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_f64_4x4x4f64 | v_mfma_f64_4x4x4f64 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_i32_16x16x16i8 | v_mfma_i32_16x16x16i8 | VOP3P | Multiply the 16x16 matrix in the first input by the 16x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_i32_16x16x32_i8 | v_mfma_i32_16x16x32_i8 | VOP3P | Multiply the 16x32 matrix in the first input by the 32x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_i32_16x16x4_4b_i8 | v_mfma_i32_16x16x4_4b_i8 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_i32_16x16x4i8 | v_mfma_i32_16x16x4i8 | VOP3P | Multiply the 16x4 matrix in the first input by the 4x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_i32_16x16x64_i8 | v_mfma_i32_16x16x64_i8 | VOP3P | Multiply the 16x64 matrix in the first input by the 64x16 matrix in the second input and add the 16x16 matrix in the third input using fused multiply… |
| v_mfma_i32_32x32x16_i8 | v_mfma_i32_32x32x16_i8 | VOP3P | Multiply the 32x16 matrix in the first input by the 16x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_i32_32x32x32_i8 | v_mfma_i32_32x32x32_i8 | VOP3P | Multiply the 32x32 matrix in the first input by the 32x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_i32_32x32x4_2b_i8 | v_mfma_i32_32x32x4_2b_i8 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_i32_32x32x4i8 | v_mfma_i32_32x32x4i8 | VOP3P | Multiply the 32x4 matrix in the first input by the 4x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_i32_32x32x8i8 | v_mfma_i32_32x32x8i8 | VOP3P | Multiply the 32x8 matrix in the first input by the 8x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_mfma_i32_4x4x4_16b_i8 | v_mfma_i32_4x4x4_16b_i8 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_i32_4x4x4i8 | v_mfma_i32_4x4x4i8 | VOP3P | Multiply the 4x4 matrix in the first input by the 4x4 matrix in the second input and add the 4x4 matrix in the third input using fused multiply add. |
| v_mfma_ld_scale_b32 | v_mfma_ld_scale_b32 | VOP3P | AMDGPU VOP3P matrix instruction operating on b32 data. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_mfma_scale_f32_16x16x128_f8f6f4 | v_mfma_scale_f32_16x16x128_f8f6f4 | VOP3P | Multiply the 16x128 matrix in the first input by the 128x16 matrix in the second input and add the 16x16 matrix in the third input using fused… |
| v_mfma_scale_f32_32x32x64_f8f6f4 | v_mfma_scale_f32_32x32x64_f8f6f4 | VOP3P | Multiply the 32x64 matrix in the first input by the 64x32 matrix in the second input and add the 32x32 matrix in the third input using fused multiply… |
| v_pk_add_bf16 | v_pk_add_bf16 | VOP3P | AMDGPU VOP3P vector instruction. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_pk_add_f16 | v_pk_add_f16 | VOP3P | Add two packed half-precision float inputs component-wise and store the result into a vector register. No carry- in or carry-out support. |
| v_pk_add_f32 | v_pk_add_f32 | VOP3P | Add two packed single-precision float inputs component-wise and store the result into a vector register. No carry-in or carry-out support. |
| v_pk_add_f64 | v_pk_add_f64 | VOP3P | AMDGPU VOP3P vector instruction operating on f64 data. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |
| v_pk_add_i16 | v_pk_add_i16 | VOP3P | Add two packed signed 16-bit integer inputs component-wise and store the result into a vector register. No carry-in or carry-out support. |
| v_pk_add_max_i16 | v_pk_add_max_i16 | VOP3P | AMDGPU VOP3P vector instruction operating on i16 data. (Format and name extracted from LLVM's AMDGPU backend source - semantics not yet curated.) |