VPKUDUS:        "vpkudus",
	VPMSUMB:        "vpmsumb",
	VPMSUMD:        "vpmsumd",
	VPMSUMH:        "vpmsumh",
	VPMSUMW:        "vpmsumw",
	VPOPCNTB:       "vpopcntb",
	VPOPCNTD:       "vpopcntd",
	VPOPCNTH:       "vpopcnth",
	VPOPCNTW:       "vpopcntw",
	VRLD:           "vrld",
	VSBOX:          "vsbox",
	VSHASIGMAD:     "vshasigmad",
	VSHASIGMAW:     "vshasigmaw",
	VSLD:           "vsld",

		v.AddArg2(y, x)
		return true
	}
}
func rewriteValuePPC64_OpCtz16(v *Value) bool {
	v_0 := v.Args[0]
	b := v.Block
	typ := &b.Func.Config.Types
	// match: (Ctz16 x)
	// result: (POPCNTW (MOVHZreg (ANDN <typ.Int16> (ADDconst <typ.Int16> [-1] x) x)))
	for {
		x := v_0
		v.reset(OpPPC64POPCNTW)
		v0 := b.NewValue0(v.Pos, OpPPC64MOVHZreg, typ.Int64)
		v1 := b.NewValue0(v.Pos, OpPPC64ANDN, typ.Int16)

	}
	return false
}
func rewriteValueS390X_OpPopCount16(v *Value) bool {
	v_0 := v.Args[0]
	b := v.Block
	typ := &b.Func.Config.Types
	// match: (PopCount16 x)
	// result: (MOVBZreg (SumBytes2 (POPCNT <typ.UInt16> x)))
	for {
		x := v_0
		v.reset(OpS390XMOVBZreg)
		v0 := b.NewValue0(v.Pos, OpS390XSumBytes2, typ.UInt8)
		v1 := b.NewValue0(v.Pos, OpS390XPOPCNT, typ.UInt16)
		v1.AddArg(x)
		v0.AddArg(v1)

// CHECK:       %[[POP_CNT:.*]] = "tf.PopulationCount"(%[[ARG_0]]) : (tensor<8xi32>) -> tensor<8xui8>
// CHECK:       %[[RES:.*]] = "tf.Cast"(%[[POP_CNT]]) <{Truncate = false}> : (tensor<8xui8>) -> tensor<8xi32>
// CHECK:       return %[[RES]]
// CHECK:         }
func.func @convert_population_count_i32(%arg0: tensor<8xi32>) -> tensor<8xi32> {
  %0 = "mhlo.popcnt"(%arg0) : (tensor<8xi32>) -> tensor<8xi32>

			bFalse := s.f.NewBlock(ssa.BlockPlain)
			bEnd := s.f.NewBlock(ssa.BlockPlain)
			b.AddEdgeTo(bTrue)
			b.AddEdgeTo(bFalse)
			b.Likely = ssa.BranchLikely // most machines have popcnt nowadays

			// We have the intrinsic - use it directly.
			s.startBlock(bTrue)
			s.vars[n] = s.newValue1(op, types.Types[types.TINT], args[0])
			s.endBlock().AddEdgeTo(bEnd)