func (t *caseTrie) lookupValue(n uint32, b byte) uint16 {
	switch {
	case n < 20:
		return uint16(caseValues[n<<6+uint32(b)])
	default:
		n -= 20
		return uint16(sparse.lookup(n, b))
	}
}

// caseValues: 22 blocks, 1408 entries, 2816 bytes
// The third block is the zero block.
var caseValues = [1408]uint16{
	// Block 0x0, offset 0x0
	0x27: 0x0054,
	0x2e: 0x0054,

	// Read resync meta header
	switch binary.LittleEndian.Uint16(data[0:2]) {
	case resyncMetaFormat:
	default:
		return brs, fmt.Errorf("resyncMeta: unknown format: %d", binary.LittleEndian.Uint16(data[0:2]))
	}
	switch binary.LittleEndian.Uint16(data[2:4]) {
	case resyncMetaVersion:
	default:
		return brs, fmt.Errorf("resyncMeta: unknown version: %d", binary.LittleEndian.Uint16(data[2:4]))
	}
	// OK, parse data.

func (t *caseTrie) lookupValue(n uint32, b byte) uint16 {
	switch {
	case n < 22:
		return uint16(caseValues[n<<6+uint32(b)])
	default:
		n -= 22
		return uint16(sparse.lookup(n, b))
	}
}

// caseValues: 24 blocks, 1536 entries, 3072 bytes
// The third block is the zero block.
var caseValues = [1536]uint16{
	// Block 0x0, offset 0x0
	0x27: 0x0054,
	0x2e: 0x0054,

		return LOP_IRR(OP_ORIS, uint32(r), REGZERO, uint32(v))
	}
	return AOP_IRR(OP_ADDIS, uint32(r), REGZERO, uint32(v))
}

func high16adjusted(d int32) uint16 {
	if d&0x8000 != 0 {
		return uint16((d >> 16) + 1)
	}
	return uint16(d >> 16)
}

func asmout(c *ctxt9, p *obj.Prog, o *Optab, out *[5]uint32) {
	o1 := uint32(0)
	o2 := uint32(0)
	o3 := uint32(0)
	o4 := uint32(0)

// is ui16. ui16 storage type for quantized type is not compatible with
// `tfl.quantize`.

func.func @uniform_quantize_op_uint16_output(%arg: tensor<2x2xf32>) -> tensor<2x2x!quant.uniform<ui16:f32, 3.000000e+0:127>> {
  %0 = stablehlo.uniform_quantize %arg : (tensor<2x2xf32>) -> tensor<2x2x!quant.uniform<ui16:f32, 3.000000e+0:127>>
  return %0 : tensor<2x2x!quant.uniform<ui16:f32, 3.000000e+0:127>>
}

  let arguments = (ins
     TFL_TensorOf<[I8, I16, I32, I64, UI8, UI16, UI32, UI64, F32, F64]>:$input,
     TFL_TensorOf<[I32]>:$dilations,
     TFL_0DTensorOf<[I8, I16, I32, I64, UI8, UI16, UI32, UI64, F32, F64]>:$padding_value
  );

  let results = (outs TFL_TensorOf<[I8, I16, I32, I64, UI8, UI16, UI32, UI64, F32, F64]>:$output);
}

def TFL_AddOp : TFL_Op<"add", [

		{name: "DIVLU", argLength: 2, reg: gp11div, typ: "(UInt32,UInt32)", asm: "DIVL", clobberFlags: true},
		{name: "DIVWU", argLength: 2, reg: gp11div, typ: "(UInt16,UInt16)", asm: "DIVW", clobberFlags: true},

		// computes -arg0, flags set for 0-arg0.
		{name: "NEGLflags", argLength: 1, reg: gp11flags, typ: "(UInt32,Flags)", asm: "NEGL", resultInArg0: true},

		cmp[i] = &xcmp[args[i]]
	}
	for i := range ops {
		op := &ops[i]
		if cmp[0][op.a1] && cmp[1][op.a2] &&
			cmp[2][op.a3] && cmp[3][op.a4] &&
			cmp[4][op.a5] && cmp[5][op.a6] {
			p.Optab = uint16(cap(optab) - cap(ops) + i + 1)
			return op
		}
	}

	// Cannot find a case; abort.
	s := ""
	for _, a := range args {
		s += fmt.Sprintf(" %v", DRconv(int(a)))
	}

(ORNshiftRO (RORconst x [c]) x [c]) => (MOVDconst [-1])

// rev16w | rev16
// ((x>>8) | (x<<8)) => (REV16W x), the type of x is uint16, "|" can also be "^" or "+".
((ADDshiftLL|ORshiftLL|XORshiftLL) <typ.UInt16> [8] (UBFX <typ.UInt16> [armBFAuxInt(8, 8)] x) x) => (REV16W x)

// ((x & 0xff00ff00)>>8) | ((x & 0x00ff00ff)<<8), "|" can also be "^" or "+".

		v.AddArg(x)
		return true
	}
	// match: (MOVHUreg (MOVWconst [c]))
	// result: (MOVWconst [int32(uint16(c))])
	for {
		if v_0.Op != OpMIPSMOVWconst {
			break
		}
		c := auxIntToInt32(v_0.AuxInt)
		v.reset(OpMIPSMOVWconst)
		v.AuxInt = int32ToAuxInt(int32(uint16(c)))
		return true
	}
	return false
}
func rewriteValueMIPS_OpMIPSMOVHload(v *Value) bool {
	v_1 := v.Args[1]