Tensor* tensor) override TF_LOCKS_EXCLUDED(mu_);

  Status MakeTensorFromProto(DeviceContext* device_context,
                             const TensorProto& tensor_proto,
                             const AllocatorAttributes alloc_attrs,
                             Tensor* tensor);

  const Metadata& metadata() { return xla_metadata_; }

  StatusOr<TaggedValue> v = tf_function.Execute(ctx_.get(), args);
  ASSERT_TRUE(v.ok()) << v.status().message();
  const TaggedValue& result = v.value();
  ExpectEquals(result.tuple()[0].tensor().get(), 2.0f);
  ExpectEquals(result.tuple()[1].tensor().get(), 4.0f);
}

TEST_P(FunctionTest, UnaryFuncCalledWithMultipleArgsFails) {
  // Construct a scalar.
  impl::TaggedValueTensor x = CreateScalarTensor<float, TF_FLOAT>(2.0f);

    new (&v.data_.dict) T(std::make_shared<T::element_type>());
    return v;
  }
  /// Constructs a TaggedValue with type TENSOR.
  static TaggedValue Tensor(tensorflow::AbstractTensorHandle* raw_ptr) {
    TaggedValue v;
    v.type_ = TENSOR;
    using T = decltype(v.data_.tensor);
    new (&v.data_.tensor) T(raw_ptr, /*add_ref=*/false);
    return v;
  }

      CHECK-LABEL: tfr.func @tf__test_identity_op(%x: !tfr.tensor) -> (!tfr.tensor) {
      CHECK-NEXT:   %[[raw_data:.*]] = tfr.quant_raw_data(%x) : (!tfr.tensor) -> (!tfr.tensor)
      CHECK-NEXT:   %[[qparam:.*]]:2 = tfr.quant_qparam(%x) : (!tfr.tensor) -> (!tfr.tensor, !tfr.tensor)
      CHECK:        %[[list:.*]] = "tfr.build_list"(%[[qparam]]#0, %[[qparam]]#0) : (!tfr.tensor, !tfr.tensor) -> !tfr.tensor_list

}

# MLIR-LABEL: func @main
# MLIR-SAME:  (%[[ARG_0:[a-z0-9]+]]: tensor<1x1x1x256x!quant.uniform<i8:f32, 0.21632751372549019:27>>) -> tensor<1x6x31x!quant.uniform<i8:f32, 0.09363494573854933:22>>
# MLIR-SAME:  control_outputs = ""
# MLIR-SAME:  inputs = "input"
# MLIR-SAME:  outputs = "output"
# MLIR:         %[[shape:.*]] = arith.constant dense<[1, -1, 31]> : tensor<3xi32>

}

// CHECK-LABEL: func @testPow
// CHECK-SAME:(%[[ARG_0:.*]]: tensor<4xf32>, %[[ARG_1:.*]]: tensor<4xf32>) -> (tensor<4xf32>, tensor<4xf32>, tensor<4xf32>)
func.func @testPow(%arg0: tensor<4xf32>, %arg1: tensor<4xf32>) -> (tensor<4xf32>, tensor<4xf32>, tensor<4xf32>) {

  %cst_zero = arith.constant dense<0.0> : tensor<f32>
  %cst_one = arith.constant dense<1.0> : tensor<f32>

    "mhlo.return"(%0, %1) : (tensor<f32>, tensor<f32>) -> ()
  },  {
    "mhlo.return"(%0, %1) : (tensor<f32>, tensor<f32>) -> ()
  // CHECK:      (tensor<i1>) -> (tensor<f32>, tensor<f32>, !mhlo.token)
  }) : (tensor<i1>) -> (tensor<f32>, tensor<f32>)

  // CHECK:      [[IF_SUBTUPLE_RESULT:%.*]] = mhlo.tuple [[IF]]#0, [[IF]]#1

  // copy-on-read mode is false.
  if (var->tensor()->RefCountIsOne()) {
    var->copy_on_read_mode.store(true);
    return absl::OkStatus();
  }
  Tensor tmp;
  if (variantType) {
    AllocatorAttributes attr;
    attr.set_on_host(true);
    TF_RETURN_IF_ERROR(context->allocate_temp(
        var->tensor()->dtype(), var->tensor()->shape(), &tmp, attr));

    const auto elements_in = var->tensor()->flat<Variant>();

// on-device shape. This is accurate for CPU and GPU devices that neither
// transpose nor pad tensors.
Status DefaultPaddedShapeFn(const Tensor& tensor, xla::Shape* shape) {
  const tensorflow::XlaTensor* xla_tensor =
      tensorflow::XlaTensor::FromTensor(&tensor);
  if (xla_tensor == nullptr) {
    return TensorShapeToXLAShape(tensor.dtype(), tensor.shape(), shape);
  }

  const xla::ShapedBuffer& shaped_buffer = xla_tensor->shaped_buffer();

// Legalize TensorList Ops iff all of them are supported.
std::unique_ptr<OperationPass<ModuleOp>> CreateLegalizeTensorListPass();

// Reduce the type precision of some tensor types if all values within that
// tensor are within the range of the reduced precision.
std::unique_ptr<OperationPass<ModuleOp>> CreateReduceTypePrecisionPass();

// Convervatively pushes transposes through elementwise ops to prepare