%0 = "tfr.cast"(%arg0) : (tensor<10x!quant.uniform<i8:f32, 0.0039133410900831223:-128>>) -> !tfr.tensor
  %1 = tfr.quant_raw_data(%0) : (!tfr.tensor) -> !tfr.tensor
  %2 = "tfr.cast"(%1) : (!tfr.tensor) -> tensor<10xi8>
  %3 = "tf.Cast"(%2) {Truncate = false} : (tensor<10xi8>) -> tensor<10xi32>
  func.return %3 : tensor<10xi32>

        var_handle_op.getContext(), var_handle_op.getSharedName()));
    var_handle_op->setLoc(loc);

    // Ex) If VarHandleOp's type is tensor<!tf_type.resource<tensor<1xf32>>>,
    // then tensor<1xf32> is the subtype.
    tensor_types.emplace_back(var_handle_op.resource_subtype());
  }

  auto builder =
      OpBuilder::atBlockTerminator(&session_init_func.getBody().front());

  %0 = "tfr.constant_tensor"(%arg0) : (vector<1xi32>) -> tensor<1xf32>
  func.return %0 : tensor<1xf32>
}

// -----

func.func @constant_tensor_invalid_3(%arg0: vector<1xi32>) -> tensor<1x1xi32> {
    // expected-error@+1 {{input and output should have same shape and element type}}
  %0 = "tfr.constant_tensor"(%arg0) : (vector<1xi32>) -> tensor<1x1xi32>
  func.return %0 : tensor<1x1xi32>
}

// -----

  %cst_1 = arith.constant dense<0> : tensor<4xi32>
  %cst_2 = arith.constant dense<0.000000e+00> : tensor<1xf32>
  %0 = "tfl.quantize"(%cst_2) {qtype = tensor<1x!quant.uniform<i32:f32, 1.0>>, volatile} : (tensor<1xf32>) -> tensor<1x!quant.uniform<i32:f32, 1.0>>
  %1 = "tfl.dequantize"(%0) : (tensor<1x!quant.uniform<i32:f32, 1.0>>) -> tensor<1xf32>
  %cst_3 = arith.constant dense<[[[[1.0]], [[1.0]], [[1.0]]]]> : tensor<1x3x1x1xf32>

    %1 = "tf.AddV2"(%0, %cst_0) {device = ""} : (tensor<?x?x?x?x3xf32>, tensor<f32>) -> tensor<?x?x?x?x3xf32>
    %2 = "tf.Floor"(%1) {device = ""} : (tensor<?x?x?x?x3xf32>) -> tensor<?x?x?x?x3xf32>
    %3 = "tf.ClipByValue"(%2, %cst_5, %cst_6) {device = ""} : (tensor<?x?x?x?x3xf32>, tensor<f32>, tensor<f32>) -> tensor<?x?x?x?x3xf32>

        unknown_rank: true
      }
    }
  }
  attr {
    key: "dtype"
    value {
      type: DT_BOOL
    }
  }
}

# CHECK-DAG: tf.IteratorGetNext{{.+}}-> (tensor<1x8xbf16>, tensor<2x?x16xf32>, tensor<*xf64>)
# CHECK-DAG: tf.IteratorGetNextSync{{.+}}-> (tensor<*xi16>, tensor<3x24xi32>, tensor<?x4x32xi64>)

  }

  // CHECK-LABEL: @body(%arg0: tensor<0xf32>) -> tensor<0xf32>
  func.func private @body(%arg0:tensor<0xf32>, %arg1:tensor<*x!tf_type.resource>) -> (tensor<0xf32>, tensor<*x!tf_type.resource>) {
    %val = "tf.ReadVariableOp"(%arg1) : (tensor<*x!tf_type.resource>) -> (tensor<0xf32>)
    %0 = "tf.Mul"(%val, %arg0) : (tensor<0xf32>, tensor<0xf32>) -> tensor<0xf32>
    func.return %0, %arg1 : tensor<0xf32>, tensor<*x!tf_type.resource>
  }

  // NOFALLBACK: XlaSvd
  %s, %u, %v = "tf.XlaSvd"(%arg0) {max_iter = 1, epsilon = 1.0E-09 : f32, precision_config = ""} : (tensor<1x1xf32>) -> (tensor<1xf32>, tensor<1x1xf32>, tensor<1x1xf32>)
  func.return %s, %u, %v : tensor<1xf32>, tensor<1x1xf32>, tensor<1x1xf32>
}

//===----------------------------------------------------------------------===//
// Random op legalizations.

  %2 = chlo.broadcast_add %1, %bias : (
      tensor<?x2x2x1xi32>, tensor<1xi32>) ->tensor<?x2x2x1xi32>
  return %2 : tensor<?x2x2x1xi32>
}

// -----

// CHECK-LABEL: func @convolution_add_add_static
func.func @convolution_add_add_static(
    %lhs: tensor<2x3x2x1xi8>, %rhs: tensor<2x1x1x1xi8>,
    %zp_offset: tensor<2x2x2x1xi32>, %bias: tensor<1xi32>
  ) -> tensor<2x2x2x1xi32> {
  // CHECK-DAG: %[[conv:.*]] = mhlo.convolution

  %7 = "tfl.add"(%2, %1) {fused_activation_function = "NONE"} : (tensor<4xf32>, tensor<  f32>) -> tensor<4xf32>
  %8 = "tfl.add"(%2, %3) {fused_activation_function = "NONE"} : (tensor<4xf32>, tensor<4xf32>) -> tensor<4xf32>
  %9 = "tfl.add"(%2, %3) {fused_activation_function = "SIGN_BIT"} : (tensor<4xf32>, tensor<4xf32>) -> tensor<4xf32>