LLVM_DEBUG(llvm::dbgs() << "Zero point value should be a tensor type. Got: "
                            << zero_points_type << ".\n");
    return failure();
  }

  if (Type zero_points_element_type = zero_points_type.getElementType();
      !mlir::isa<IntegerType>(zero_points_element_type)) {
    LLVM_DEBUG(llvm::dbgs() << "Zero point should be an integer type. Got: "
                            << zero_points_element_type << ".\n");

        llvm::dyn_cast<TF::ConstOp>(fold_results[0].getDefiningOp()).getValue(),
        &tensor));
    xla::Shape xla_shape;
    TF_RETURN_IF_ERROR(tensorflow::TensorShapeToXLAShape(
        tensor.dtype(), tensor.shape(), &xla_shape));
    xla::PjRtClient::HostBufferSemantics host_buffer_semantics =
        xla::PjRtClient::HostBufferSemantics::kImmutableUntilTransferCompletes;
    TF_ASSIGN_OR_RETURN(
        auto buffer,

ElementsAttr GetTensorValueAsElementsAttr(
    TF::VarHandleOp var_handle_op, const tensorflow::Tensor& resource_tensor,
    const tensorflow::DeviceMgr* mgr, OpBuilder builder) {
  if (resource_tensor.dtype() != tensorflow::DT_RESOURCE) {
    return GetTensorValueAsElementsAttr(resource_tensor, builder);
  }

  auto handle = resource_tensor.scalar<tensorflow::ResourceHandle>()();

}

// Return true if the passed quantized type is unsigned.
bool QuantizedTypeIsUnsigned(Type type) {
  return TypeSwitch<Type, bool>(type)
      .Case<mlir::TF::Qint8Type>([](Type) { return false; })
      .Case<mlir::TF::Qint16Type>([](Type) { return false; })
      .Case<mlir::TF::Qint32Type>([](Type) { return false; })
      .Case<mlir::TF::Quint8Type>([](Type) { return true; })

    } else {
      return tensorflow::Status(absl::StatusCode::kInvalidArgument,
                                "Unsupported type");
    }
  } else if (auto itype = mlir::dyn_cast<mlir::IntegerType>(element_type)) {
    if (element_type.isSignedInteger()) {
      switch (itype.getWidth()) {
        case 8:
          return DenseElementsAttr::get<int8_t>(shaped_type,

    Location loc = graph_func.getLoc();
    FunctionType ftype = graph_func.getFunctionType();

    func::FuncOp func = rewriter.create<func::FuncOp>(
        graph_func.getLoc(),
        graph_func->getAttrOfType<StringAttr>(SymbolTable::getSymbolAttrName())
            .getValue(),
        ftype);

    func->setAttrs(graph_func->getAttrs());

    llvm::SmallVector<Type> arg_types;
    llvm::SmallVector<Type> res_types;

string AttrView::DefaultValue() const {
  const AttrValue &attr_value = attr_.default_value();
  switch (attr_value.value_case()) {
    case AttrValue::VALUE_NOT_SET:
      return "";
    case AttrValue::kType:
      return DataType_Name(attr_value.type());
    case AttrValue::kS:
      return "\"" + attr_value.s() + "\"";
    case AttrValue::kI:
      return std::to_string(attr_value.i());
    case AttrValue::kF:

                    Attribute value) {
  // Handle the case where the type and value are already tensors.
  if (mlir::isa<TensorType>(type) && mlir::isa<ElementsAttr>(value)) {
    result.addTypes(type);
    result.addAttribute("value", value);
    return;
  }

  // Otherwise, default to the attribute builder.
  ConstOp::build(builder, result, value);

      assert(while_caller.body_function() == func);
      while_callers.push_back(while_caller);
    }
  }
  return while_callers;
}

bool IsResourceType(Type type) {
  return mlir::isa<TF::ResourceType>(getElementTypeOrSelf(type));
}

bool OnlyOperatesOnCompositeDevices(
    TF::TPUExecuteAndUpdateVariablesOp& op,
    const TF::SideEffectAnalysis::Info& side_effect_analysis,

    }
    if (ftype.isF64()) {
      return tflite::TensorType_COMPLEX128;
    }
    return Status(absl::StatusCode::kInvalidArgument, "Unsupported type");
  } else if (auto itype = mlir::dyn_cast<mlir::IntegerType>(type)) {
    switch (itype.getWidth()) {
      case 1:
        return tflite::TensorType_BOOL;
      case 4:
        if (itype.isUnsigned()) {
          return Status(absl::StatusCode::kInvalidArgument,