code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

//     opcode name,                                         return type,    arg1 type,      arg2 type,      arg3 type,      arg4 type,      arg4 type,      ...
OPCODE(Phi,                                                 Opaque,                                                                                )
OPCODE(Identity,                                            Opaque,         Opaque,                                                                         )
OPCODE(Void,                                                Void,                                                                                )
OPCODE(ConditionRef,                                        U1,             U1,                                                                             )
OPCODE(Reference,                                           Void,           Opaque,                                                                         )
OPCODE(PhiMove,                                             Void,           Opaque,         Opaque,                                                         )

// Special operations
OPCODE(Prologue,                                            Void,                                                                                )
OPCODE(Epilogue,                                            Void,                                                                                )
OPCODE(Join,                                                Void,                                                                                )
OPCODE(DemoteToHelperInvocation,                            Void,                                                                                )
OPCODE(EmitVertex,                                          Void,           U32,                                                                            )
OPCODE(EndPrimitive,                                        Void,           U32,                                                                            )

// Barriers
OPCODE(Barrier,                                             Void,                                                                                )
OPCODE(WorkgroupMemoryBarrier,                              Void,                                                                                )
OPCODE(DeviceMemoryBarrier,                                 Void,                                                                                )

// Context getters/setters
OPCODE(GetRegister,                                         U32,            Reg,                                                                            )
OPCODE(SetRegister,                                         Void,           Reg,            U32,                                                            )
OPCODE(GetPred,                                             U1,             Pred,                                                                           )
OPCODE(SetPred,                                             Void,           Pred,           U1,                                                             )
OPCODE(GetGotoVariable,                                     U1,             U32,                                                                            )
OPCODE(SetGotoVariable,                                     Void,           U32,            U1,                                                             )
OPCODE(GetIndirectBranchVariable,                           U32,                                                                                )
OPCODE(SetIndirectBranchVariable,                           Void,           U32,                                                                            )
OPCODE(GetCbufU8,                                           U32,            U32,            U32,                                                            )
OPCODE(GetCbufS8,                                           U32,            U32,            U32,                                                            )
OPCODE(GetCbufU16,                                          U32,            U32,            U32,                                                            )
OPCODE(GetCbufS16,                                          U32,            U32,            U32,                                                            )
OPCODE(GetCbufU32,                                          U32,            U32,            U32,                                                            )
OPCODE(GetCbufF32,                                          F32,            U32,            U32,                                                            )
OPCODE(GetCbufU32x2,                                        U32x2,          U32,            U32,                                                            )
OPCODE(GetAttribute,                                        F32,            Attribute,      U32,                                                            )
OPCODE(GetAttributeU32,                                     U32,            Attribute,      U32,                                                            )
OPCODE(SetAttribute,                                        Void,           Attribute,      F32,            U32,                                            )
OPCODE(GetAttributeIndexed,                                 F32,            U32,            U32,                                                            )
OPCODE(SetAttributeIndexed,                                 Void,           U32,            F32,            U32,                                            )
OPCODE(GetPatch,                                            F32,            Patch,                                                                          )
OPCODE(SetPatch,                                            Void,           Patch,          F32,                                                            )
OPCODE(SetFragColor,                                        Void,           U32,            U32,            F32,                                            )
OPCODE(SetSampleMask,                                       Void,           U32,                                                                            )
OPCODE(SetFragDepth,                                        Void,           F32,                                                                            )
OPCODE(GetZFlag,                                            U1,             Void,                                                                           )
OPCODE(GetSFlag,                                            U1,             Void,                                                                           )
OPCODE(GetCFlag,                                            U1,             Void,                                                                           )
OPCODE(GetOFlag,                                            U1,             Void,                                                                           )
OPCODE(SetZFlag,                                            Void,           U1,                                                                             )
OPCODE(SetSFlag,                                            Void,           U1,                                                                             )
OPCODE(SetCFlag,                                            Void,           U1,                                                                             )
OPCODE(SetOFlag,                                            Void,           U1,                                                                             )
OPCODE(WorkgroupId,                                         U32x3,                                                                                )
OPCODE(LocalInvocationId,                                   U32x3,                                                                                )
OPCODE(InvocationId,                                        U32,                                                                                )
OPCODE(InvocationInfo,                                      U32,                                                                                )
OPCODE(SampleId,                                            U32,                                                                                )
OPCODE(IsHelperInvocation,                                  U1,                                                                                )
OPCODE(SR_WScaleFactorXY,                                   U32,                                                                                )
OPCODE(SR_WScaleFactorZ,                                    U32,                                                                                )
OPCODE(YDirection,                                          F32,                                                                                )
OPCODE(ResolutionDownFactor,                                F32,                                                                                )
OPCODE(RenderArea,                                          F32x4,                                                                                )

// Undefined
OPCODE(UndefU1,                                             U1,                                                                                )
OPCODE(UndefU8,                                             U8,                                                                                )
OPCODE(UndefU16,                                            U16,                                                                                )
OPCODE(UndefU32,                                            U32,                                                                                )
OPCODE(UndefU64,                                            U64,                                                                                )

// Memory operations
OPCODE(LoadGlobalU8,                                        U32,            Opaque,                                                                         )
OPCODE(LoadGlobalS8,                                        U32,            Opaque,                                                                         )
OPCODE(LoadGlobalU16,                                       U32,            Opaque,                                                                         )
OPCODE(LoadGlobalS16,                                       U32,            Opaque,                                                                         )
OPCODE(LoadGlobal32,                                        U32,            Opaque,                                                                         )
OPCODE(LoadGlobal64,                                        U32x2,          Opaque,                                                                         )
OPCODE(LoadGlobal128,                                       U32x4,          Opaque,                                                                         )
OPCODE(WriteGlobalU8,                                       Void,           Opaque,         U32,                                                            )
OPCODE(WriteGlobalS8,                                       Void,           Opaque,         U32,                                                            )
OPCODE(WriteGlobalU16,                                      Void,           Opaque,         U32,                                                            )
OPCODE(WriteGlobalS16,                                      Void,           Opaque,         U32,                                                            )
OPCODE(WriteGlobal32,                                       Void,           Opaque,         U32,                                                            )
OPCODE(WriteGlobal64,                                       Void,           Opaque,         U32x2,                                                          )
OPCODE(WriteGlobal128,                                      Void,           Opaque,         U32x4,                                                          )

// Storage buffer operations
OPCODE(LoadStorageU8,                                       U32,            U32,            U32,                                                            )
OPCODE(LoadStorageS8,                                       U32,            U32,            U32,                                                            )
OPCODE(LoadStorageU16,                                      U32,            U32,            U32,                                                            )
OPCODE(LoadStorageS16,                                      U32,            U32,            U32,                                                            )
OPCODE(LoadStorage32,                                       U32,            U32,            U32,                                                            )
OPCODE(LoadStorage64,                                       U32x2,          U32,            U32,                                                            )
OPCODE(LoadStorage128,                                      U32x4,          U32,            U32,                                                            )
OPCODE(WriteStorageU8,                                      Void,           U32,            U32,            U32,                                            )
OPCODE(WriteStorageS8,                                      Void,           U32,            U32,            U32,                                            )
OPCODE(WriteStorageU16,                                     Void,           U32,            U32,            U32,                                            )
OPCODE(WriteStorageS16,                                     Void,           U32,            U32,            U32,                                            )
OPCODE(WriteStorage32,                                      Void,           U32,            U32,            U32,                                            )
OPCODE(WriteStorage64,                                      Void,           U32,            U32,            U32x2,                                          )
OPCODE(WriteStorage128,                                     Void,           U32,            U32,            U32x4,                                          )

// Local memory operations
OPCODE(LoadLocal,                                           U32,            U32,                                                                            )
OPCODE(WriteLocal,                                          Void,           U32,            U32,                                                            )

// Shared memory operations
OPCODE(LoadSharedU8,                                        U32,            U32,                                                                            )
OPCODE(LoadSharedS8,                                        U32,            U32,                                                                            )
OPCODE(LoadSharedU16,                                       U32,            U32,                                                                            )
OPCODE(LoadSharedS16,                                       U32,            U32,                                                                            )
OPCODE(LoadSharedU32,                                       U32,            U32,                                                                            )
OPCODE(LoadSharedU64,                                       U32x2,          U32,                                                                            )
OPCODE(LoadSharedU128,                                      U32x4,          U32,                                                                            )
OPCODE(WriteSharedU8,                                       Void,           U32,            U32,                                                            )
OPCODE(WriteSharedU16,                                      Void,           U32,            U32,                                                            )
OPCODE(WriteSharedU32,                                      Void,           U32,            U32,                                                            )
OPCODE(WriteSharedU64,                                      Void,           U32,            U32x2,                                                          )
OPCODE(WriteSharedU128,                                     Void,           U32,            U32x4,                                                          )

// Vector utility
OPCODE(CompositeConstructU32x2,                             U32x2,          U32,            U32,                                                            )
OPCODE(CompositeConstructU32x3,                             U32x3,          U32,            U32,            U32,                                            )
OPCODE(CompositeConstructU32x4,                             U32x4,          U32,            U32,            U32,            U32,                            )
OPCODE(CompositeExtractU32x2,                               U32,            U32x2,          U32,                                                            )
OPCODE(CompositeExtractU32x3,                               U32,            U32x3,          U32,                                                            )
OPCODE(CompositeExtractU32x4,                               U32,            U32x4,          U32,                                                            )
OPCODE(CompositeInsertU32x2,                                U32x2,          U32x2,          U32,            U32,                                            )
OPCODE(CompositeInsertU32x3,                                U32x3,          U32x3,          U32,            U32,                                            )
OPCODE(CompositeInsertU32x4,                                U32x4,          U32x4,          U32,            U32,                                            )
OPCODE(CompositeConstructF16x2,                             F16x2,          F16,            F16,                                                            )
OPCODE(CompositeConstructF16x3,                             F16x3,          F16,            F16,            F16,                                            )
OPCODE(CompositeConstructF16x4,                             F16x4,          F16,            F16,            F16,            F16,                            )
OPCODE(CompositeExtractF16x2,                               F16,            F16x2,          U32,                                                            )
OPCODE(CompositeExtractF16x3,                               F16,            F16x3,          U32,                                                            )
OPCODE(CompositeExtractF16x4,                               F16,            F16x4,          U32,                                                            )
OPCODE(CompositeInsertF16x2,                                F16x2,          F16x2,          F16,            U32,                                            )
OPCODE(CompositeInsertF16x3,                                F16x3,          F16x3,          F16,            U32,                                            )
OPCODE(CompositeInsertF16x4,                                F16x4,          F16x4,          F16,            U32,                                            )
OPCODE(CompositeConstructF32x2,                             F32x2,          F32,            F32,                                                            )
OPCODE(CompositeConstructF32x3,                             F32x3,          F32,            F32,            F32,                                            )
OPCODE(CompositeConstructF32x4,                             F32x4,          F32,            F32,            F32,            F32,                            )
OPCODE(CompositeExtractF32x2,                               F32,            F32x2,          U32,                                                            )
OPCODE(CompositeExtractF32x3,                               F32,            F32x3,          U32,                                                            )
OPCODE(CompositeExtractF32x4,                               F32,            F32x4,          U32,                                                            )
OPCODE(CompositeInsertF32x2,                                F32x2,          F32x2,          F32,            U32,                                            )
OPCODE(CompositeInsertF32x3,                                F32x3,          F32x3,          F32,            U32,                                            )
OPCODE(CompositeInsertF32x4,                                F32x4,          F32x4,          F32,            U32,                                            )
OPCODE(CompositeConstructF64x2,                             F64x2,          F64,            F64,                                                            )
OPCODE(CompositeConstructF64x3,                             F64x3,          F64,            F64,            F64,                                            )
OPCODE(CompositeConstructF64x4,                             F64x4,          F64,            F64,            F64,            F64,                            )
OPCODE(CompositeExtractF64x2,                               F64,            F64x2,          U32,                                                            )
OPCODE(CompositeExtractF64x3,                               F64,            F64x3,          U32,                                                            )
OPCODE(CompositeExtractF64x4,                               F64,            F64x4,          U32,                                                            )
OPCODE(CompositeInsertF64x2,                                F64x2,          F64x2,          F64,            U32,                                            )
OPCODE(CompositeInsertF64x3,                                F64x3,          F64x3,          F64,            U32,                                            )
OPCODE(CompositeInsertF64x4,                                F64x4,          F64x4,          F64,            U32,                                            )

// Select operations
OPCODE(SelectU1,                                            U1,             U1,             U1,             U1,                                             )
OPCODE(SelectU8,                                            U8,             U1,             U8,             U8,                                             )
OPCODE(SelectU16,                                           U16,            U1,             U16,            U16,                                            )
OPCODE(SelectU32,                                           U32,            U1,             U32,            U32,                                            )
OPCODE(SelectU64,                                           U64,            U1,             U64,            U64,                                            )
OPCODE(SelectF16,                                           F16,            U1,             F16,            F16,                                            )
OPCODE(SelectF32,                                           F32,            U1,             F32,            F32,                                            )
OPCODE(SelectF64,                                           F64,            U1,             F64,            F64,                                            )

// Bitwise conversions
OPCODE(BitCastU16F16,                                       U16,            F16,                                                                            )
OPCODE(BitCastU32F32,                                       U32,            F32,                                                                            )
OPCODE(BitCastU64F64,                                       U64,            F64,                                                                            )
OPCODE(BitCastF16U16,                                       F16,            U16,                                                                            )
OPCODE(BitCastF32U32,                                       F32,            U32,                                                                            )
OPCODE(BitCastF64U64,                                       F64,            U64,                                                                            )
OPCODE(PackUint2x32,                                        U64,            U32x2,                                                                          )
OPCODE(UnpackUint2x32,                                      U32x2,          U64,                                                                            )
OPCODE(PackFloat2x16,                                       U32,            F16x2,                                                                          )
OPCODE(UnpackFloat2x16,                                     F16x2,          U32,                                                                            )
OPCODE(PackHalf2x16,                                        U32,            F32x2,                                                                          )
OPCODE(UnpackHalf2x16,                                      F32x2,          U32,                                                                            )
OPCODE(PackDouble2x32,                                      F64,            U32x2,                                                                          )
OPCODE(UnpackDouble2x32,                                    U32x2,          F64,                                                                            )

// Pseudo-operation, handled specially at final emit
OPCODE(GetZeroFromOp,                                       U1,             Opaque,                                                                         )
OPCODE(GetSignFromOp,                                       U1,             Opaque,                                                                         )
OPCODE(GetCarryFromOp,                                      U1,             Opaque,                                                                         )
OPCODE(GetOverflowFromOp,                                   U1,             Opaque,                                                                         )
OPCODE(GetSparseFromOp,                                     U1,             Opaque,                                                                         )
OPCODE(GetInBoundsFromOp,                                   U1,             Opaque,                                                                         )

// Floating-point operations
OPCODE(FPAbs16,                                             F16,            F16,                                                                            )
OPCODE(FPAbs32,                                             F32,            F32,                                                                            )
OPCODE(FPAbs64,                                             F64,            F64,                                                                            )
OPCODE(FPAdd16,                                             F16,            F16,            F16,                                                            )
OPCODE(FPAdd32,                                             F32,            F32,            F32,                                                            )
OPCODE(FPAdd64,                                             F64,            F64,            F64,                                                            )
OPCODE(FPFma16,                                             F16,            F16,            F16,            F16,                                            )
OPCODE(FPFma32,                                             F32,            F32,            F32,            F32,                                            )
OPCODE(FPFma64,                                             F64,            F64,            F64,            F64,                                            )
OPCODE(FPMax32,                                             F32,            F32,            F32,                                                            )
OPCODE(FPMax64,                                             F64,            F64,            F64,                                                            )
OPCODE(FPMin32,                                             F32,            F32,            F32,                                                            )
OPCODE(FPMin64,                                             F64,            F64,            F64,                                                            )
OPCODE(FPMul16,                                             F16,            F16,            F16,                                                            )
OPCODE(FPMul32,                                             F32,            F32,            F32,                                                            )
OPCODE(FPMul64,                                             F64,            F64,            F64,                                                            )
OPCODE(FPNeg16,                                             F16,            F16,                                                                            )
OPCODE(FPNeg32,                                             F32,            F32,                                                                            )
OPCODE(FPNeg64,                                             F64,            F64,                                                                            )
OPCODE(FPRecip32,                                           F32,            F32,                                                                            )
OPCODE(FPRecip64,                                           F64,            F64,                                                                            )
OPCODE(FPRecipSqrt32,                                       F32,            F32,                                                                            )
OPCODE(FPRecipSqrt64,                                       F64,            F64,                                                                            )
OPCODE(FPSqrt,                                              F32,            F32,                                                                            )
OPCODE(FPSin,                                               F32,            F32,                                                                            )
OPCODE(FPExp2,                                              F32,            F32,                                                                            )
OPCODE(FPCos,                                               F32,            F32,                                                                            )
OPCODE(FPLog2,                                              F32,            F32,                                                                            )
OPCODE(FPSaturate16,                                        F16,            F16,                                                                            )
OPCODE(FPSaturate32,                                        F32,            F32,                                                                            )
OPCODE(FPSaturate64,                                        F64,            F64,                                                                            )
OPCODE(FPClamp16,                                           F16,            F16,            F16,            F16,                                            )
OPCODE(FPClamp32,                                           F32,            F32,            F32,            F32,                                            )
OPCODE(FPClamp64,                                           F64,            F64,            F64,            F64,                                            )
OPCODE(FPRoundEven16,                                       F16,            F16,                                                                            )
OPCODE(FPRoundEven32,                                       F32,            F32,                                                                            )
OPCODE(FPRoundEven64,                                       F64,            F64,                                                                            )
OPCODE(FPFloor16,                                           F16,            F16,                                                                            )
OPCODE(FPFloor32,                                           F32,            F32,                                                                            )
OPCODE(FPFloor64,                                           F64,            F64,                                                                            )
OPCODE(FPCeil16,                                            F16,            F16,                                                                            )
OPCODE(FPCeil32,                                            F32,            F32,                                                                            )
OPCODE(FPCeil64,                                            F64,            F64,                                                                            )
OPCODE(FPTrunc16,                                           F16,            F16,                                                                            )
OPCODE(FPTrunc32,                                           F32,            F32,                                                                            )
OPCODE(FPTrunc64,                                           F64,            F64,                                                                            )

OPCODE(FPOrdEqual16,                                        U1,             F16,            F16,                                                            )
OPCODE(FPOrdEqual32,                                        U1,             F32,            F32,                                                            )
OPCODE(FPOrdEqual64,                                        U1,             F64,            F64,                                                            )
OPCODE(FPUnordEqual16,                                      U1,             F16,            F16,                                                            )
OPCODE(FPUnordEqual32,                                      U1,             F32,            F32,                                                            )
OPCODE(FPUnordEqual64,                                      U1,             F64,            F64,                                                            )
OPCODE(FPOrdNotEqual16,                                     U1,             F16,            F16,                                                            )
OPCODE(FPOrdNotEqual32,                                     U1,             F32,            F32,                                                            )
OPCODE(FPOrdNotEqual64,                                     U1,             F64,            F64,                                                            )
OPCODE(FPUnordNotEqual16,                                   U1,             F16,            F16,                                                            )
OPCODE(FPUnordNotEqual32,                                   U1,             F32,            F32,                                                            )
OPCODE(FPUnordNotEqual64,                                   U1,             F64,            F64,                                                            )
OPCODE(FPOrdLessThan16,                                     U1,             F16,            F16,                                                            )
OPCODE(FPOrdLessThan32,                                     U1,             F32,            F32,                                                            )
OPCODE(FPOrdLessThan64,                                     U1,             F64,            F64,                                                            )
OPCODE(FPUnordLessThan16,                                   U1,             F16,            F16,                                                            )
OPCODE(FPUnordLessThan32,                                   U1,             F32,            F32,                                                            )
OPCODE(FPUnordLessThan64,                                   U1,             F64,            F64,                                                            )
OPCODE(FPOrdGreaterThan16,                                  U1,             F16,            F16,                                                            )
OPCODE(FPOrdGreaterThan32,                                  U1,             F32,            F32,                                                            )
OPCODE(FPOrdGreaterThan64,                                  U1,             F64,            F64,                                                            )
OPCODE(FPUnordGreaterThan16,                                U1,             F16,            F16,                                                            )
OPCODE(FPUnordGreaterThan32,                                U1,             F32,            F32,                                                            )
OPCODE(FPUnordGreaterThan64,                                U1,             F64,            F64,                                                            )
OPCODE(FPOrdLessThanEqual16,                                U1,             F16,            F16,                                                            )
OPCODE(FPOrdLessThanEqual32,                                U1,             F32,            F32,                                                            )
OPCODE(FPOrdLessThanEqual64,                                U1,             F64,            F64,                                                            )
OPCODE(FPUnordLessThanEqual16,                              U1,             F16,            F16,                                                            )
OPCODE(FPUnordLessThanEqual32,                              U1,             F32,            F32,                                                            )
OPCODE(FPUnordLessThanEqual64,                              U1,             F64,            F64,                                                            )
OPCODE(FPOrdGreaterThanEqual16,                             U1,             F16,            F16,                                                            )
OPCODE(FPOrdGreaterThanEqual32,                             U1,             F32,            F32,                                                            )
OPCODE(FPOrdGreaterThanEqual64,                             U1,             F64,            F64,                                                            )
OPCODE(FPUnordGreaterThanEqual16,                           U1,             F16,            F16,                                                            )
OPCODE(FPUnordGreaterThanEqual32,                           U1,             F32,            F32,                                                            )
OPCODE(FPUnordGreaterThanEqual64,                           U1,             F64,            F64,                                                            )
OPCODE(FPIsNan16,                                           U1,             F16,                                                                            )
OPCODE(FPIsNan32,                                           U1,             F32,                                                                            )
OPCODE(FPIsNan64,                                           U1,             F64,                                                                            )

// Integer operations
OPCODE(IAdd32,                                              U32,            U32,            U32,                                                            )
OPCODE(IAdd64,                                              U64,            U64,            U64,                                                            )
OPCODE(ISub32,                                              U32,            U32,            U32,                                                            )
OPCODE(ISub64,                                              U64,            U64,            U64,                                                            )
OPCODE(IMul32,                                              U32,            U32,            U32,                                                            )
OPCODE(SDiv32,                                              U32,            U32,            U32,                                                            )
OPCODE(UDiv32,                                              U32,            U32,            U32,                                                            )
OPCODE(INeg32,                                              U32,            U32,                                                                            )
OPCODE(INeg64,                                              U64,            U64,                                                                            )
OPCODE(IAbs32,                                              U32,            U32,                                                                            )
OPCODE(ShiftLeftLogical32,                                  U32,            U32,            U32,                                                            )
OPCODE(ShiftLeftLogical64,                                  U64,            U64,            U32,                                                            )
OPCODE(ShiftRightLogical32,                                 U32,            U32,            U32,                                                            )
OPCODE(ShiftRightLogical64,                                 U64,            U64,            U32,                                                            )
OPCODE(ShiftRightArithmetic32,                              U32,            U32,            U32,                                                            )
OPCODE(ShiftRightArithmetic64,                              U64,            U64,            U32,                                                            )
OPCODE(BitwiseAnd32,                                        U32,            U32,            U32,                                                            )
OPCODE(BitwiseOr32,                                         U32,            U32,            U32,                                                            )
OPCODE(BitwiseXor32,                                        U32,            U32,            U32,                                                            )
OPCODE(BitFieldInsert,                                      U32,            U32,            U32,            U32,            U32,                            )
OPCODE(BitFieldSExtract,                                    U32,            U32,            U32,            U32,                                            )
OPCODE(BitFieldUExtract,                                    U32,            U32,            U32,            U32,                                            )
OPCODE(BitReverse32,                                        U32,            U32,                                                                            )
OPCODE(BitCount32,                                          U32,            U32,                                                                            )
OPCODE(BitwiseNot32,                                        U32,            U32,                                                                            )

OPCODE(FindSMsb32,                                          U32,            U32,                                                                            )
OPCODE(FindUMsb32,                                          U32,            U32,                                                                            )
OPCODE(SMin32,                                              U32,            U32,            U32,                                                            )
OPCODE(UMin32,                                              U32,            U32,            U32,                                                            )
OPCODE(SMax32,                                              U32,            U32,            U32,                                                            )
OPCODE(UMax32,                                              U32,            U32,            U32,                                                            )
OPCODE(SClamp32,                                            U32,            U32,            U32,            U32,                                            )
OPCODE(UClamp32,                                            U32,            U32,            U32,            U32,                                            )
OPCODE(SLessThan,                                           U1,             U32,            U32,                                                            )
OPCODE(ULessThan,                                           U1,             U32,            U32,                                                            )
OPCODE(IEqual,                                              U1,             U32,            U32,                                                            )
OPCODE(SLessThanEqual,                                      U1,             U32,            U32,                                                            )
OPCODE(ULessThanEqual,                                      U1,             U32,            U32,                                                            )
OPCODE(SGreaterThan,                                        U1,             U32,            U32,                                                            )
OPCODE(UGreaterThan,                                        U1,             U32,            U32,                                                            )
OPCODE(INotEqual,                                           U1,             U32,            U32,                                                            )
OPCODE(SGreaterThanEqual,                                   U1,             U32,            U32,                                                            )
OPCODE(UGreaterThanEqual,                                   U1,             U32,            U32,                                                            )

// Atomic operations
OPCODE(SharedAtomicIAdd32,                                  U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicSMin32,                                  U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicUMin32,                                  U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicSMax32,                                  U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicUMax32,                                  U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicInc32,                                   U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicDec32,                                   U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicAnd32,                                   U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicOr32,                                    U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicXor32,                                   U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicExchange32,                              U32,            U32,            U32,                                                            )
OPCODE(SharedAtomicExchange64,                              U64,            U32,            U64,                                                            )
OPCODE(SharedAtomicExchange32x2,                            U32x2,          U32,            U32x2,                                                          )

OPCODE(GlobalAtomicIAdd32,                                  U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicSMin32,                                  U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicUMin32,                                  U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicSMax32,                                  U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicUMax32,                                  U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicInc32,                                   U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicDec32,                                   U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicAnd32,                                   U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicOr32,                                    U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicXor32,                                   U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicExchange32,                              U32,            U64,            U32,                                                            )
OPCODE(GlobalAtomicIAdd64,                                  U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicSMin64,                                  U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicUMin64,                                  U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicSMax64,                                  U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicUMax64,                                  U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicAnd64,                                   U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicOr64,                                    U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicXor64,                                   U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicExchange64,                              U64,            U64,            U64,                                                            )
OPCODE(GlobalAtomicIAdd32x2,                                U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicSMin32x2,                                U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicUMin32x2,                                U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicSMax32x2,                                U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicUMax32x2,                                U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicAnd32x2,                                 U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicOr32x2,                                  U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicXor32x2,                                 U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicExchange32x2,                            U32x2,          U32x2,          U32x2,                                                          )
OPCODE(GlobalAtomicAddF32,                                  F32,            U64,            F32,                                                            )
OPCODE(GlobalAtomicAddF16x2,                                U32,            U64,            F16x2,                                                          )
OPCODE(GlobalAtomicAddF32x2,                                U32,            U64,            F32x2,                                                          )
OPCODE(GlobalAtomicMinF16x2,                                U32,            U64,            F16x2,                                                          )
OPCODE(GlobalAtomicMinF32x2,                                U32,            U64,            F32x2,                                                          )
OPCODE(GlobalAtomicMaxF16x2,                                U32,            U64,            F16x2,                                                          )
OPCODE(GlobalAtomicMaxF32x2,                                U32,            U64,            F32x2,                                                          )

OPCODE(StorageAtomicIAdd32,                                 U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicSMin32,                                 U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicUMin32,                                 U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicSMax32,                                 U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicUMax32,                                 U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicInc32,                                  U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicDec32,                                  U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicAnd32,                                  U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicOr32,                                   U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicXor32,                                  U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicExchange32,                             U32,            U32,            U32,            U32,                                            )
OPCODE(StorageAtomicIAdd64,                                 U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicSMin64,                                 U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicUMin64,                                 U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicSMax64,                                 U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicUMax64,                                 U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicAnd64,                                  U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicOr64,                                   U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicXor64,                                  U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicExchange64,                             U64,            U32,            U32,            U64,                                            )
OPCODE(StorageAtomicIAdd32x2,                               U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicSMin32x2,                               U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicUMin32x2,                               U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicSMax32x2,                               U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicUMax32x2,                               U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicAnd32x2,                                U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicOr32x2,                                 U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicXor32x2,                                U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicExchange32x2,                           U32x2,          U32,            U32,            U32x2,                                          )
OPCODE(StorageAtomicAddF32,                                 F32,            U32,            U32,            F32,                                            )
OPCODE(StorageAtomicAddF16x2,                               U32,            U32,            U32,            F16x2,                                          )
OPCODE(StorageAtomicAddF32x2,                               U32,            U32,            U32,            F32x2,                                          )
OPCODE(StorageAtomicMinF16x2,                               U32,            U32,            U32,            F16x2,                                          )
OPCODE(StorageAtomicMinF32x2,                               U32,            U32,            U32,            F32x2,                                          )
OPCODE(StorageAtomicMaxF16x2,                               U32,            U32,            U32,            F16x2,                                          )
OPCODE(StorageAtomicMaxF32x2,                               U32,            U32,            U32,            F32x2,                                          )

// Logical operations
OPCODE(LogicalOr,                                           U1,             U1,             U1,                                                             )
OPCODE(LogicalAnd,                                          U1,             U1,             U1,                                                             )
OPCODE(LogicalXor,                                          U1,             U1,             U1,                                                             )
OPCODE(LogicalNot,                                          U1,             U1,                                                                             )

// Conversion operations
OPCODE(ConvertS16F16,                                       U32,            F16,                                                                            )
OPCODE(ConvertS16F32,                                       U32,            F32,                                                                            )
OPCODE(ConvertS16F64,                                       U32,            F64,                                                                            )
OPCODE(ConvertS32F16,                                       U32,            F16,                                                                            )
OPCODE(ConvertS32F32,                                       U32,            F32,                                                                            )
OPCODE(ConvertS32F64,                                       U32,            F64,                                                                            )
OPCODE(ConvertS64F16,                                       U64,            F16,                                                                            )
OPCODE(ConvertS64F32,                                       U64,            F32,                                                                            )
OPCODE(ConvertS64F64,                                       U64,            F64,                                                                            )
OPCODE(ConvertU16F16,                                       U32,            F16,                                                                            )
OPCODE(ConvertU16F32,                                       U32,            F32,                                                                            )
OPCODE(ConvertU16F64,                                       U32,            F64,                                                                            )
OPCODE(ConvertU32F16,                                       U32,            F16,                                                                            )
OPCODE(ConvertU32F32,                                       U32,            F32,                                                                            )
OPCODE(ConvertU32F64,                                       U32,            F64,                                                                            )
OPCODE(ConvertU64F16,                                       U64,            F16,                                                                            )
OPCODE(ConvertU64F32,                                       U64,            F32,                                                                            )
OPCODE(ConvertU64F64,                                       U64,            F64,                                                                            )
OPCODE(ConvertU64U32,                                       U64,            U32,                                                                            )
OPCODE(ConvertU32U64,                                       U32,            U64,                                                                            )
OPCODE(ConvertF16F32,                                       F16,            F32,                                                                            )
OPCODE(ConvertF32F16,                                       F32,            F16,                                                                            )
OPCODE(ConvertF32F64,                                       F32,            F64,                                                                            )
OPCODE(ConvertF64F32,                                       F64,            F32,                                                                            )
OPCODE(ConvertF16S8,                                        F16,            U32,                                                                            )
OPCODE(ConvertF16S16,                                       F16,            U32,                                                                            )
OPCODE(ConvertF16S32,                                       F16,            U32,                                                                            )
OPCODE(ConvertF16S64,                                       F16,            U64,                                                                            )
OPCODE(ConvertF16U8,                                        F16,            U32,                                                                            )
OPCODE(ConvertF16U16,                                       F16,            U32,                                                                            )
OPCODE(ConvertF16U32,                                       F16,            U32,                                                                            )
OPCODE(ConvertF16U64,                                       F16,            U64,                                                                            )
OPCODE(ConvertF32S8,                                        F32,            U32,                                                                            )
OPCODE(ConvertF32S16,                                       F32,            U32,                                                                            )
OPCODE(ConvertF32S32,                                       F32,            U32,                                                                            )
OPCODE(ConvertF32S64,                                       F32,            U64,                                                                            )
OPCODE(ConvertF32U8,                                        F32,            U32,                                                                            )
OPCODE(ConvertF32U16,                                       F32,            U32,                                                                            )
OPCODE(ConvertF32U32,                                       F32,            U32,                                                                            )
OPCODE(ConvertF32U64,                                       F32,            U64,                                                                            )
OPCODE(ConvertF64S8,                                        F64,            U32,                                                                            )
OPCODE(ConvertF64S16,                                       F64,            U32,                                                                            )
OPCODE(ConvertF64S32,                                       F64,            U32,                                                                            )
OPCODE(ConvertF64S64,                                       F64,            U64,                                                                            )
OPCODE(ConvertF64U8,                                        F64,            U32,                                                                            )
OPCODE(ConvertF64U16,                                       F64,            U32,                                                                            )
OPCODE(ConvertF64U32,                                       F64,            U32,                                                                            )
OPCODE(ConvertF64U64,                                       F64,            U64,                                                                            )

// Image operations
OPCODE(BindlessImageSampleImplicitLod,                      F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BindlessImageSampleExplicitLod,                      F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BindlessImageSampleDrefImplicitLod,                  F32,            U32,            Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(BindlessImageSampleDrefExplicitLod,                  F32,            U32,            Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(BindlessImageGather,                                 F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BindlessImageGatherDref,                             F32x4,          U32,            Opaque,         Opaque,         Opaque,         F32,            )
OPCODE(BindlessImageFetch,                                  F32x4,          U32,            Opaque,         Opaque,         U32,            Opaque,         )
OPCODE(BindlessImageQueryDimensions,                        U32x4,          U32,            U32,            U1,                                             )
OPCODE(BindlessImageQueryLod,                               F32x4,          U32,            Opaque,                                                         )
OPCODE(BindlessImageGradient,                               F32x4,          U32,            Opaque,         Opaque,         Opaque,         Opaque,         )
OPCODE(BindlessImageRead,                                   U32x4,          U32,            Opaque,                                                         )
OPCODE(BindlessImageWrite,                                  Void,           U32,            Opaque,         U32x4,                                          )

OPCODE(BoundImageSampleImplicitLod,                         F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BoundImageSampleExplicitLod,                         F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BoundImageSampleDrefImplicitLod,                     F32,            U32,            Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(BoundImageSampleDrefExplicitLod,                     F32,            U32,            Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(BoundImageGather,                                    F32x4,          U32,            Opaque,         Opaque,         Opaque,                         )
OPCODE(BoundImageGatherDref,                                F32x4,          U32,            Opaque,         Opaque,         Opaque,         F32,            )
OPCODE(BoundImageFetch,                                     F32x4,          U32,            Opaque,         Opaque,         U32,            Opaque,         )
OPCODE(BoundImageQueryDimensions,                           U32x4,          U32,            U32,            U1,                                             )
OPCODE(BoundImageQueryLod,                                  F32x4,          U32,            Opaque,                                                         )
OPCODE(BoundImageGradient,                                  F32x4,          U32,            Opaque,         Opaque,         Opaque,         Opaque,         )
OPCODE(BoundImageRead,                                      U32x4,          U32,            Opaque,                                                         )
OPCODE(BoundImageWrite,                                     Void,           U32,            Opaque,         U32x4,                                          )

OPCODE(ImageSampleImplicitLod,                              F32x4,          Opaque,         Opaque,         Opaque,         Opaque,                         )
OPCODE(ImageSampleExplicitLod,                              F32x4,          Opaque,         Opaque,         Opaque,         Opaque,                         )
OPCODE(ImageSampleDrefImplicitLod,                          F32,            Opaque,         Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(ImageSampleDrefExplicitLod,                          F32,            Opaque,         Opaque,         F32,            Opaque,         Opaque,         )
OPCODE(ImageGather,                                         F32x4,          Opaque,         Opaque,         Opaque,         Opaque,                         )
OPCODE(ImageGatherDref,                                     F32x4,          Opaque,         Opaque,         Opaque,         Opaque,         F32,            )
OPCODE(ImageFetch,                                          F32x4,          Opaque,         Opaque,         Opaque,         U32,            Opaque,         )
OPCODE(ImageQueryDimensions,                                U32x4,          Opaque,         U32,            U1,                                             )
OPCODE(ImageQueryLod,                                       F32x4,          Opaque,         Opaque,                                                         )
OPCODE(ImageGradient,                                       F32x4,          Opaque,         Opaque,         Opaque,         Opaque,         Opaque,         )
OPCODE(ImageRead,                                           U32x4,          Opaque,         Opaque,                                                         )
OPCODE(ImageWrite,                                          Void,           Opaque,         Opaque,         U32x4,                                          )

OPCODE(IsTextureScaled,                                     U1,             U32,                                                                            )
OPCODE(IsImageScaled,                                       U1,             U32,                                                                            )

// Atomic Image operations

OPCODE(BindlessImageAtomicIAdd32,                           U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicSMin32,                           U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicUMin32,                           U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicSMax32,                           U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicUMax32,                           U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicInc32,                            U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicDec32,                            U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicAnd32,                            U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicOr32,                             U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicXor32,                            U32,            U32,            Opaque,            U32,                                         )
OPCODE(BindlessImageAtomicExchange32,                       U32,            U32,            Opaque,            U32,                                         )

OPCODE(BoundImageAtomicIAdd32,                              U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicSMin32,                              U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicUMin32,                              U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicSMax32,                              U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicUMax32,                              U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicInc32,                               U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicDec32,                               U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicAnd32,                               U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicOr32,                                U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicXor32,                               U32,            U32,            Opaque,            U32,                                         )
OPCODE(BoundImageAtomicExchange32,                          U32,            U32,            Opaque,            U32,                                         )

OPCODE(ImageAtomicIAdd32,                                   U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicSMin32,                                   U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicUMin32,                                   U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicSMax32,                                   U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicUMax32,                                   U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicInc32,                                    U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicDec32,                                    U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicAnd32,                                    U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicOr32,                                     U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicXor32,                                    U32,            Opaque,            Opaque,            U32,                                      )
OPCODE(ImageAtomicExchange32,                               U32,            Opaque,            Opaque,            U32,                                      )

// Warp operations
OPCODE(LaneId,                                              U32,                                                                                )
OPCODE(VoteAll,                                             U1,             U1,                                                                             )
OPCODE(VoteAny,                                             U1,             U1,                                                                             )
OPCODE(VoteEqual,                                           U1,             U1,                                                                             )
OPCODE(SubgroupBallot,                                      U32,            U1,                                                                             )
OPCODE(SubgroupEqMask,                                      U32,                                                                                )
OPCODE(SubgroupLtMask,                                      U32,                                                                                )
OPCODE(SubgroupLeMask,                                      U32,                                                                                )
OPCODE(SubgroupGtMask,                                      U32,                                                                                )
OPCODE(SubgroupGeMask,                                      U32,                                                                                )
OPCODE(ShuffleIndex,                                        U32,            U32,            U32,            U32,            U32,                            )
OPCODE(ShuffleUp,                                           U32,            U32,            U32,            U32,            U32,                            )
OPCODE(ShuffleDown,                                         U32,            U32,            U32,            U32,            U32,                            )
OPCODE(ShuffleButterfly,                                    U32,            U32,            U32,            U32,            U32,                            )
OPCODE(FSwizzleAdd,                                         F32,            F32,            F32,            U32,                                            )
OPCODE(DPdxFine,                                            F32,            F32,                                                                            )
OPCODE(DPdyFine,                                            F32,            F32,                                                                            )
OPCODE(DPdxCoarse,                                          F32,            F32,                                                                            )
OPCODE(DPdyCoarse,                                          F32,            F32,                                                                            )


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#include "shader_recompiler/exception.h"
#include "shader_recompiler/frontend/ir/patch.h"

namespace Shader::IR {

bool IsGeneric(Patch patch) noexcept {
    return patch >= Patch::Component0 && patch <= Patch::Component119;
}

u32 GenericPatchIndex(Patch patch) {
    if (!IsGeneric(patch)) {
        throw InvalidArgument("Patch {} is not generic", patch);
    }
    return (static_cast<u32>(patch) - static_cast<u32>(Patch::Component0)) / 4;
}

u32 GenericPatchElement(Patch patch) {
    if (!IsGeneric(patch)) {
        throw InvalidArgument("Patch {} is not generic", patch);
    }
    return (static_cast<u32>(patch) - static_cast<u32>(Patch::Component0)) % 4;
}

} // namespace Shader::IR

repo_name: litucks/torzu

license: mit


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#pragma once

#include "common/common_types.h"

namespace Shader::IR {

enum class Patch : u64 {
    TessellationLodLeft,
    TessellationLodTop,
    TessellationLodRight,
    TessellationLodBottom,
    TessellationLodInteriorU,
    TessellationLodInteriorV,
    Component0,
    Component1,
    Component2,
    Component3,
    Component4,
    Component5,
    Component6,
    Component7,
    Component8,
    Component9,
    Component10,
    Component11,
    Component12,
    Component13,
    Component14,
    Component15,
    Component16,
    Component17,
    Component18,
    Component19,
    Component20,
    Component21,
    Component22,
    Component23,
    Component24,
    Component25,
    Component26,
    Component27,
    Component28,
    Component29,
    Component30,
    Component31,
    Component32,
    Component33,
    Component34,
    Component35,
    Component36,
    Component37,
    Component38,
    Component39,
    Component40,
    Component41,
    Component42,
    Component43,
    Component44,
    Component45,
    Component46,
    Component47,
    Component48,
    Component49,
    Component50,
    Component51,
    Component52,
    Component53,
    Component54,
    Component55,
    Component56,
    Component57,
    Component58,
    Component59,
    Component60,
    Component61,
    Component62,
    Component63,
    Component64,
    Component65,
    Component66,
    Component67,
    Component68,
    Component69,
    Component70,
    Component71,
    Component72,
    Component73,
    Component74,
    Component75,
    Component76,
    Component77,
    Component78,
    Component79,
    Component80,
    Component81,
    Component82,
    Component83,
    Component84,
    Component85,
    Component86,
    Component87,
    Component88,
    Component89,
    Component90,
    Component91,
    Component92,
    Component93,
    Component94,
    Component95,
    Component96,
    Component97,
    Component98,
    Component99,
    Component100,
    Component101,
    Component102,
    Component103,
    Component104,
    Component105,
    Component106,
    Component107,
    Component108,
    Component109,
    Component110,
    Component111,
    Component112,
    Component113,
    Component114,
    Component115,
    Component116,
    Component117,
    Component118,
    Component119,
};
static_assert(static_cast<u64>(Patch::Component119) == 125);

[[nodiscard]] bool IsGeneric(Patch patch) noexcept;

[[nodiscard]] u32 GenericPatchIndex(Patch patch);

[[nodiscard]] u32 GenericPatchElement(Patch patch);

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#include <algorithm>

#include <boost/container/flat_set.hpp>
#include <boost/container/small_vector.hpp>

#include "shader_recompiler/frontend/ir/basic_block.h"
#include "shader_recompiler/frontend/ir/post_order.h"

namespace Shader::IR {

BlockList PostOrder(const AbstractSyntaxNode& root) {
    boost::container::small_vector<Block*, 16> block_stack;
    boost::container::flat_set<Block*> visited;
    BlockList post_order_blocks;

    if (root.type != AbstractSyntaxNode::Type::Block) {
        throw LogicError("First node in abstract syntax list root is not a block");
    }
    Block* const first_block{root.data.block};
    visited.insert(first_block);
    block_stack.push_back(first_block);

    while (!block_stack.empty()) {
        Block* const block{block_stack.back()};
        const auto visit{[&](Block* branch) {
            if (!visited.insert(branch).second) {
                return false;
            }
            // Calling push_back twice is faster than insert on MSVC
            block_stack.push_back(block);
            block_stack.push_back(branch);
            return true;
        }};
        block_stack.pop_back();
        if (std::ranges::none_of(block->ImmSuccessors(), visit)) {
            post_order_blocks.push_back(block);
        }
    }
    return post_order_blocks;
}

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#pragma once

#include "shader_recompiler/frontend/ir/abstract_syntax_list.h"
#include "shader_recompiler/frontend/ir/basic_block.h"

namespace Shader::IR {

BlockList PostOrder(const AbstractSyntaxNode& root);

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#pragma once

#include <fmt/format.h>

namespace Shader::IR {

enum class Pred : u64 {
    P0,
    P1,
    P2,
    P3,
    P4,
    P5,
    P6,
    PT,
};

constexpr size_t NUM_USER_PREDS = 7;
constexpr size_t NUM_PREDS = 8;

[[nodiscard]] constexpr size_t PredIndex(Pred pred) noexcept {
    return static_cast<size_t>(pred);
}

} // namespace Shader::IR

template <>
struct fmt::formatter<Shader::IR::Pred> {
    constexpr auto parse(format_parse_context& ctx) {
        return ctx.begin();
    }
    template <typename FormatContext>
    auto format(const Shader::IR::Pred& pred, FormatContext& ctx) {
        if (pred == Shader::IR::Pred::PT) {
            return fmt::format_to(ctx.out(), "PT");
        } else {
            return fmt::format_to(ctx.out(), "P{}", static_cast<int>(pred));
        }
    }
};


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#include <map>
#include <string>

#include <fmt/format.h>

#include "shader_recompiler/frontend/ir/basic_block.h"
#include "shader_recompiler/frontend/ir/program.h"
#include "shader_recompiler/frontend/ir/value.h"

namespace Shader::IR {

std::string DumpProgram(const Program& program) {
    size_t index{0};
    std::map<const IR::Inst*, size_t> inst_to_index;
    std::map<const IR::Block*, size_t> block_to_index;

    for (const IR::Block* const block : program.blocks) {
        block_to_index.emplace(block, index);
        ++index;
    }
    std::string ret;
    for (const auto& block : program.blocks) {
        ret += IR::DumpBlock(*block, block_to_index, inst_to_index, index) + '\n';
    }
    return ret;
}

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#pragma once

#include <array>
#include <string>

#include "shader_recompiler/frontend/ir/abstract_syntax_list.h"
#include "shader_recompiler/frontend/ir/basic_block.h"
#include "shader_recompiler/program_header.h"
#include "shader_recompiler/shader_info.h"
#include "shader_recompiler/stage.h"

namespace Shader::IR {

struct Program {
    AbstractSyntaxList syntax_list;
    BlockList blocks;
    BlockList post_order_blocks;
    Info info;
    Stage stage{};
    std::array<u32, 3> workgroup_size{};
    OutputTopology output_topology{};
    u32 output_vertices{};
    u32 invocations{};
    u32 local_memory_size{};
    u32 shared_memory_size{};
    bool is_geometry_passthrough{};
};

[[nodiscard]] std::string DumpProgram(const Program& program);

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#include <array>
#include <string>

#include "shader_recompiler/frontend/ir/type.h"

namespace Shader::IR {

std::string NameOf(Type type) {
    static constexpr std::array names{
        "Opaque", "Label", "Reg",   "Pred",  "Attribute", "U1",    "U8",    "U16",   "U32",
        "U64",    "F16",   "F32",   "F64",   "U32x2",     "U32x3", "U32x4", "F16x2", "F16x3",
        "F16x4",  "F32x2", "F32x3", "F32x4", "F64x2",     "F64x3", "F64x4",
    };
    const size_t bits{static_cast<size_t>(type)};
    if (bits == 0) {
        return "Void";
    }
    std::string result;
    for (size_t i = 0; i < names.size(); i++) {
        if ((bits & (size_t{1} << i)) != 0) {
            if (!result.empty()) {
                result += '|';
            }
            result += names[i];
        }
    }
    return result;
}

bool AreTypesCompatible(Type lhs, Type rhs) noexcept {
    return lhs == rhs || lhs == Type::Opaque || rhs == Type::Opaque;
}

} // namespace Shader::IR


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#pragma once

#include <string>

#include <fmt/format.h>

#include "common/common_funcs.h"
#include "shader_recompiler/exception.h"

namespace Shader::IR {

enum class Type {
    Void = 0,
    Opaque = 1 << 0,
    Reg = 1 << 1,
    Pred = 1 << 2,
    Attribute = 1 << 3,
    Patch = 1 << 4,
    U1 = 1 << 5,
    U8 = 1 << 6,
    U16 = 1 << 7,
    U32 = 1 << 8,
    U64 = 1 << 9,
    F16 = 1 << 10,
    F32 = 1 << 11,
    F64 = 1 << 12,
    U32x2 = 1 << 13,
    U32x3 = 1 << 14,
    U32x4 = 1 << 15,
    F16x2 = 1 << 16,
    F16x3 = 1 << 17,
    F16x4 = 1 << 18,
    F32x2 = 1 << 19,
    F32x3 = 1 << 20,
    F32x4 = 1 << 21,
    F64x2 = 1 << 22,
    F64x3 = 1 << 23,
    F64x4 = 1 << 24,
};
DECLARE_ENUM_FLAG_OPERATORS(Type)

[[nodiscard]] std::string NameOf(Type type);

[[nodiscard]] bool AreTypesCompatible(Type lhs, Type rhs) noexcept;

} // namespace Shader::IR

template <>
struct fmt::formatter<Shader::IR::Type> {
    constexpr auto parse(format_parse_context& ctx) {
        return ctx.begin();
    }
    template <typename FormatContext>
    auto format(const Shader::IR::Type& type, FormatContext& ctx) {
        return fmt::format_to(ctx.out(), "{}", NameOf(type));
    }
};


code:
// SPDX-FileCopyrightText: Copyright 2021 yuzu Emulator Project
// SPDX-License-Identifier: GPL-2.0-or-later

#include "shader_recompiler/frontend/ir/value.h"

namespace Shader::IR {

Value::Value(IR::Inst* value) noexcept : type{Type::Opaque}, inst{value} {}

Value::Value(IR::Reg value) noexcept : type{Type::Reg}, reg{value} {}

Value::Value(IR::Pred value) noexcept : type{Type::Pred}, pred{value} {}

Value::Value(IR::Attribute value) noexcept : type{Type::Attribute}, attribute{value} {}

Value::Value(IR::Patch value) noexcept : type{Type::Patch}, patch{value} {}

Value::Value(bool value) noexcept : type{Type::U1}, imm_u1{value} {}

Value::Value(u8 value) noexcept : type{Type::U8}, imm_u8{value} {}

Value::Value(u16 value) noexcept : type{Type::U16}, imm_u16{value} {}

Value::Value(u32 value) noexcept : type{Type::U32}, imm_u32{value} {}

Value::Value(f32 value) noexcept : type{Type::F32}, imm_f32{value} {}

Value::Value(u64 value) noexcept : type{Type::U64}, imm_u64{value} {}

Value::Value(f64 value) noexcept : type{Type::F64}, imm_f64{value} {}

IR::Type Value::Type() const noexcept {
    if (IsPhi()) {
        // The type of a phi node is stored in its flags
        return inst->Flags<IR::Type>();
    }
    if (IsIdentity()) {
        return inst->Arg(0).Type();
    }
    if (type == Type::Opaque) {
        return inst->Type();
    }
    return type;
}

bool Value::operator==(const Value& other) const {
    if (type != other.type) {
        return false;
    }
    switch (type) {
    case Type::Void:
        return true;
    case Type::Opaque:
        return inst == other.inst;
    case Type::Reg:
        return reg == other.reg;
    case Type::Pred:
        return pred == other.pred;
    case Type::Attribute:
        return attribute == other.attribute;
    case Type::Patch:
        return patch == other.patch;
    case Type::U1:
        return imm_u1 == other.imm_u1;
    case Type::U8:
        return imm_u8 == other.imm_u8;
    case Type::U16:
    case Type::F16:
        return imm_u16 == other.imm_u16;
    case Type::U32:
    case Type::F32:
        return imm_u32 == other.imm_u32;
    case Type::U64:
    case Type::F64:
        return imm_u64 == other.imm_u64;
    case Type::U32x2:
    case Type::U32x3:
    case Type::U32x4:
    case Type::F16x2:
    case Type::F16x3:
    case Type::F16x4:
    case Type::F32x2:
    case Type::F32x3:
    case Type::F32x4:
    case Type::F64x2:
    case Type::F64x3:
    case Type::F64x4:
        break;
    }
    throw LogicError("Invalid type {}", type);
}

bool Value::operator!=(const Value& other) const {
    return !operator==(other);
}

} // namespace Shader::IR