// CHECK-NEXT: mhlo.or
  %0 = "tf.BitwiseOr"(%arg0, %arg1) : (tensor<4xui32>, tensor<4xui32>) -> tensor<4xui32>
  func.return %0: tensor<4xui32>
}

// CHECK-LABEL: func @bitwise_xor
func.func @bitwise_xor(%arg0: tensor<4xi32>, %arg1: tensor<4xi32>) -> tensor<4xi32> {
  // CHECK-NEXT: mhlo.xor
  %0 = "tf.BitwiseXor"(%arg0, %arg1) : (tensor<4xi32>, tensor<4xi32>) -> tensor<4xi32>
  func.return %0: tensor<4xi32>
}

  RELU_0_TO_1 = 152,
  UNSORTED_SEGMENT_PROD = 153,
  UNSORTED_SEGMENT_MAX = 154,
  UNSORTED_SEGMENT_SUM = 155,
  ATAN2 = 156,
  UNSORTED_SEGMENT_MIN = 157,
  SIGN = 158,
  BITCAST = 159,
  BITWISE_XOR = 160,
  RIGHT_SHIFT = 161,
  // All Operators start with STABLEHLO_ prefixes are subject to change
  // Many of the ops below can not be executed by TFlite runtime

}

// -----

// CHECK-LABEL: testBitwiseXor
func.func @testBitwiseXor(%arg0: tensor<8xui32>, %arg1: tensor<8xui32>) -> tensor<8xui32> {
  // CHECK: "tfl.bitwise_xor"(%arg0, %arg1)
  %0 = "tfl.bitwise_xor"(%arg0, %arg1) : (tensor<8xui32>, tensor<8xui32>) -> tensor<8xui32>
  func.return %0 : tensor<8xui32>
  // CHECK: return %0 : tensor<8xui32>
}

// -----

  %0 = "chlo.broadcast_or"(%arg0, %arg1) : (tensor<?xi32>, tensor<1xi32>) -> tensor<?xi32>
  func.return %0 : tensor<?xi32>
}

// CHECK-LABEL:   func @bitwise_xor(
// CHECK-SAME:                      %[[VAL_0:.*]]: tensor<4xi32>,
// CHECK-SAME:                      %[[VAL_1:.*]]: tensor<4xi32>) -> tensor<4xi32> {

  %0 = "tf.BitwiseXor"(%arg0, %arg1) : (tensor<8xui32>, tensor<8xui32>) -> tensor<8xui32>
  func.return %0 : tensor<8xui32>

  // CHECK-LABEL: testBitwiseXor
  // CHECK: %[[RES0:.*]] = "tfl.bitwise_xor"(%arg0, %arg1) : (tensor<8xui32>, tensor<8xui32>) -> tensor<8xui32>
  // CHECK: return %[[RES0]] : tensor<8xui32>
}

func.func @testRightShift(%arg0: tensor<8xui32>, %arg1: tensor<8xui32>) -> tensor<8xui32> {

  // from the TfLiteTensors.
  let hasOptions = 0;

  let hasFolder = 1;

  let hasVerifier = 1;
}

def TFL_BitwiseXorOp : TFL_Op<"bitwise_xor", [
    Commutative,
    SameOperandsAndResultElementType,
    Pure]> {
  let summary = "Bitwise Xor operator";

  let description = [{
    Elementwise computes the bitwise XOR of `lhs` and `rhs`.

    *   Added 8-bit and 16-bit support for `floor_div` and `floor_mod`.
    *   Added 16-bit and 32-bit int support for the built-in op `bitcast`.
    *   Added 8-bit/16-bit/32-bit int/uint support for the built-in op `bitwise_xor`
    *   Added int16 indices support for built-in op `gather` and `gather_nd`.
    *   Added 8-bit/16-bit/32-bit int/uint support for the built-in op `right_shift`

    "RELU_0_TO_1",
    "UNSORTED_SEGMENT_PROD",
    "UNSORTED_SEGMENT_MAX",
    "UNSORTED_SEGMENT_SUM",
    "ATAN2",
    "UNSORTED_SEGMENT_MIN",
    "SIGN",
    "BITCAST",
    "BITWISE_XOR",
    "RIGHT_SHIFT",
    "STABLEHLO_LOGISTIC",
    "STABLEHLO_ADD",
    "STABLEHLO_DIVIDE",
    "STABLEHLO_MULTIPLY",
    "STABLEHLO_MAXIMUM",
    "STABLEHLO_RESHAPE",
    "STABLEHLO_CLAMP",

for dtype in dtype_list:
  lhs = tf.constant([0, 5, 3, 14], dtype=dtype)
  rhs = tf.constant([5, 0, 7, 11], dtype=dtype)
  exp = tf.constant([5, 5, 4, 5],  dtype=tf.float32)

  res = bitwise_ops.bitwise_xor(lhs, rhs)
  tf.assert_equal(tf.cast(res, tf.float32), exp) # TRUE
```
  }];

  let arguments = (ins
    TF_IntTensor:$x,
    TF_IntTensor:$y
  );

  let results = (outs