diff --git a/src/pcms/coupler/field_communicator.hpp b/src/pcms/coupler/field_communicator.hpp
index 3d1014a0..b1874507 100644
--- a/src/pcms/coupler/field_communicator.hpp
+++ b/src/pcms/coupler/field_communicator.hpp
@@ -35,7 +35,11 @@ class FieldCommunicator
       serializer_(std::move(serializer))
   {
     PCMS_ALWAYS_ASSERT(&layout_comm.GetLayout() == &field.GetLayout());
-    comm_buffer_.resize(layout_comm.GetMsgSize());
+    // The exchange plan is per DOF holder; the field payload carries
+    // num_components values per holder, so the buffer is scaled accordingly.
+    comm_buffer_.resize(
+      layout_comm.GetMsgSize() *
+      static_cast<size_t>(field.GetLayout().GetNumComponents()));
     comm_ =
       layout_comm_.GetChannel().CreateComm<T>(name, layout_comm.GetMPIComm());
     layout_comm_.SetOutMessageLayout<T>(comm_);
diff --git a/src/pcms/coupler/field_exchange_planner.cpp b/src/pcms/coupler/field_exchange_planner.cpp
index 83ed0eb5..4bdcd574 100644
--- a/src/pcms/coupler/field_exchange_planner.cpp
+++ b/src/pcms/coupler/field_exchange_planner.cpp
@@ -180,7 +180,7 @@ static ReversePartitionMap2 BuildReversePartitionMap(
   auto owned = layout.GetOwnedHost();
   auto class_dims = layout.GetDOFHolderClassificationDimensionsHost();
   auto class_ids = layout.GetDOFHolderClassificationIdsHost();
-  auto coords = layout.GetDOFHolderCoordinates().GetCoordinates();
+  auto coords = layout.GetDOFHolderCoordinates().GetValues();
   auto ent_offsets = layout.GetEntOffsets();
   int mesh_dim = static_cast<int>(coords.extent(1));
 
diff --git a/src/pcms/coupler/field_layout_communicator.h b/src/pcms/coupler/field_layout_communicator.h
index bc0179f3..68d3572b 100644
--- a/src/pcms/coupler/field_layout_communicator.h
+++ b/src/pcms/coupler/field_layout_communicator.h
@@ -41,7 +41,19 @@ class FieldLayoutCommunicator
   template <typename T>
   void SetOutMessageLayout(redev::BidirectionalComm<T>& comm)
   {
-    comm.SetOutMessageLayout(plan_.dest_ranks, plan_.offsets);
+    // plan_.offsets is a per-holder CSR layout. Each holder carries
+    // num_components field values, so the field message's offsets are scaled
+    // by num_components while the destination ranks are unchanged.
+    const int num_comp = layout_.GetNumComponents();
+    if (num_comp == 1) {
+      comm.SetOutMessageLayout(plan_.dest_ranks, plan_.offsets);
+      return;
+    }
+    redev::LOs scaled_offsets(plan_.offsets.size());
+    for (size_t i = 0; i < plan_.offsets.size(); ++i) {
+      scaled_offsets[i] = plan_.offsets[i] * num_comp;
+    }
+    comm.SetOutMessageLayout(plan_.dest_ranks, scaled_offsets);
   }
 
   void UpdateLayout();
diff --git a/src/pcms/coupler/field_serializer.h b/src/pcms/coupler/field_serializer.h
index fba5af57..87246456 100644
--- a/src/pcms/coupler/field_serializer.h
+++ b/src/pcms/coupler/field_serializer.h
@@ -20,13 +20,22 @@ class FieldSerializer
   {
     auto data = field.GetDOFHolderDataHost();
     auto owned = layout.GetOwnedHost();
+    // The exchange plan is per DOF holder: owned[i] and permutation[i] are
+    // indexed by holder. All num_components components of a holder share its
+    // location, so they occupy one contiguous block permutation[i]*num_comp in
+    // the wire buffer.
     if (buffer.size() > 0) {
-      for (LO i = 0; i < static_cast<LO>(data.size()); ++i) {
-        if (owned[i])
-          buffer[permutation[i]] = data[i];
+      const LO num_dof = static_cast<LO>(data.extent(0));
+      const LO num_comp = static_cast<LO>(data.extent(1));
+      for (LO i = 0; i < num_dof; ++i) {
+        if (owned[i]) {
+          for (LO c = 0; c < num_comp; ++c) {
+            buffer[permutation[i] * num_comp + c] = data(i, c);
+          }
+        }
       }
     }
-    return static_cast<int>(data.size());
+    return static_cast<int>(buffer.size());
   }
 
   virtual void Deserialize(
@@ -34,13 +43,19 @@ class FieldSerializer
     Rank1View<const T, HostMemorySpace> buffer,
     Rank1View<const LO, HostMemorySpace> permutation) const
   {
-    Kokkos::View<T*, HostMemorySpace> sorted("sorted", permutation.size());
+    const LO num_dof = layout.GetNumOwnedDofHolder();
+    const LO num_comp = layout.GetNumComponents();
+    Kokkos::View<T*, HostMemorySpace> sorted("sorted", layout.OwnedSize());
     auto owned = layout.GetOwnedHost();
-    for (LO i = 0; i < static_cast<LO>(sorted.size()); ++i) {
-      if (owned[i])
-        sorted[i] = buffer[permutation[i]];
+    for (LO i = 0; i < num_dof; ++i) {
+      if (owned[i]) {
+        for (LO c = 0; c < num_comp; ++c) {
+          sorted[i * num_comp + c] = buffer[permutation[i] * num_comp + c];
+        }
+      }
     }
-    field.SetDOFHolderDataHost(make_const_array_view(sorted));
+    field.SetDOFHolderDataHost(
+      Rank2View<const T, HostMemorySpace>(sorted.data(), num_dof, num_comp));
   }
 
   virtual ~FieldSerializer() noexcept = default;
diff --git a/src/pcms/coupler/serializer/xgc.h b/src/pcms/coupler/serializer/xgc.h
index 8a6f6cfe..ce09cf74 100644
--- a/src/pcms/coupler/serializer/xgc.h
+++ b/src/pcms/coupler/serializer/xgc.h
@@ -35,10 +35,16 @@ class XGCFieldSerializer : public FieldSerializer<T>
 
     auto data = xgc_field->GetDOFHolderDataHost();
     auto owned = layout.GetOwnedHost();
+    // Per-holder plan: owned[i]/permutation[i] index holders; a holder's
+    // num_components values form one contiguous block in the wire buffer.
     if (buffer.size() > 0) {
-      for (LO i = 0; i < static_cast<LO>(data.size()); ++i) {
+      const LO num_dof = static_cast<LO>(data.extent(0));
+      const LO num_comp = static_cast<LO>(data.extent(1));
+      for (LO i = 0; i < num_dof; ++i) {
         if (owned[i]) {
-          buffer[permutation[i]] = data[i];
+          for (LO c = 0; c < num_comp; ++c) {
+            buffer[permutation[i] * num_comp + c] = data(i, c);
+          }
         }
       }
     }
@@ -58,14 +64,20 @@ class XGCFieldSerializer : public FieldSerializer<T>
 
     auto current = xgc_field->GetDOFHolderDataHost();
     auto owned = layout.GetOwnedHost();
+    const LO num_dof = static_cast<LO>(current.extent(0));
+    const LO num_comp = static_cast<LO>(current.extent(1));
     std::vector<T> full_data(current.size());
-    for (size_t i = 0; i < current.size(); ++i) {
-      full_data[i] = current[i];
+    for (LO i = 0; i < num_dof; ++i) {
+      for (LO c = 0; c < num_comp; ++c) {
+        full_data[i * num_comp + c] = current(i, c);
+      }
     }
     if (rank_participates_) {
-      for (LO i = 0; i < static_cast<LO>(current.size()); ++i) {
+      for (LO i = 0; i < num_dof; ++i) {
         if (owned[i]) {
-          full_data[i] = buffer[permutation[i]];
+          for (LO c = 0; c < num_comp; ++c) {
+            full_data[i * num_comp + c] = buffer[permutation[i] * num_comp + c];
+          }
         }
       }
     }
@@ -73,8 +85,9 @@ class XGCFieldSerializer : public FieldSerializer<T>
     MPI_Bcast(full_data.data(), static_cast<int>(full_data.size()),
               pcms::GetMPIType(T{}), 0, plane_comm_);
 
-    xgc_field->SetDOFHolderDataHost(
-      Rank1View<const T, HostMemorySpace>(full_data.data(), full_data.size()));
+    xgc_field->SetDOFHolderDataHost(Rank2View<const T, HostMemorySpace>(
+      full_data.data(), layout.GetNumOwnedDofHolder(),
+      layout.GetNumComponents()));
   }
 
 private:
diff --git a/src/pcms/discretization/discretization/omega_h.hpp b/src/pcms/discretization/discretization/omega_h.hpp
index deeeab90..3fcb7eab 100644
--- a/src/pcms/discretization/discretization/omega_h.hpp
+++ b/src/pcms/discretization/discretization/omega_h.hpp
@@ -28,6 +28,9 @@ class OmegaHDiscretization : public Discretization
   Rank1View<const ClassificationId, DeviceMemorySpace>
   GetEntityClassificationIds(int entity_dim) const override;
 
+  // The Omega_h mesh this discretization is defined on.
+  Omega_h::Mesh& GetMesh() const noexcept { return mesh_; }
+
 private:
   static constexpr int max_entity_dim_ = Region;
 
diff --git a/src/pcms/field/coordinate_system.h b/src/pcms/field/coordinate_system.h
index 2ff6f5a1..ba627cd2 100644
--- a/src/pcms/field/coordinate_system.h
+++ b/src/pcms/field/coordinate_system.h
@@ -32,8 +32,8 @@ class CoordinateView
     return coordinate_system_;
   }
 
-  [[nodiscard]] Rank2View<const Real, MemorySpace, LayoutPolicy>
-  GetCoordinates() const noexcept
+  [[nodiscard]] Rank2View<const Real, MemorySpace, LayoutPolicy> GetValues()
+    const noexcept
   {
     return coordinates_;
   }
@@ -42,7 +42,7 @@ class CoordinateView
   // CoordinateTransformation as they are unsafe (i.e., you can break class
   // invariant) passkey pattern?
   [[nodiscard]] Rank2View<const Real, MemorySpace, LayoutPolicy>
-  GetCoordinates() noexcept
+  GetValues() noexcept
   {
     return coordinates_;
   }
diff --git a/src/pcms/field/data/mesh_fields.h b/src/pcms/field/data/mesh_fields.h
index 5d243c1f..7e5857e8 100644
--- a/src/pcms/field/data/mesh_fields.h
+++ b/src/pcms/field/data/mesh_fields.h
@@ -36,31 +36,35 @@ class MeshFieldsFieldData : public FieldData<T>
 
   const FieldMetadata& GetMetadata() const override { return metadata_; }
 
-  Rank1View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
+  Rank2View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
   {
     Kokkos::deep_copy(host_data_, device_data_);
-    return make_const_array_view(host_data_);
+    return Rank2View<const T, HostMemorySpace>(host_data_.data(),
+                                               layout_->GetNumOwnedDofHolder(),
+                                               layout_->GetNumComponents());
   }
 
-  void SetDOFHolderDataHost(Rank1View<const T, HostMemorySpace> values) override
+  void SetDOFHolderDataHost(Rank2View<const T, HostMemorySpace> values) override
   {
     PCMS_ALWAYS_ASSERT(values.size() ==
                        static_cast<size_t>(layout_->OwnedSize()));
-    CopyHostRank1ViewToDeviceView(device_data_, values);
-    SyncBackend(make_const_array_view(device_data_));
+    CopyHostRank2ViewToDeviceView(device_data_, values);
+    SyncBackend(GetDOFHolderData());
   }
 
-  Rank1View<const T, DeviceMemorySpace> GetDOFHolderData() const override
+  Rank2View<const T, DeviceMemorySpace> GetDOFHolderData() const override
   {
-    return make_const_array_view(device_data_);
+    return Rank2View<const T, DeviceMemorySpace>(
+      device_data_.data(), layout_->GetNumOwnedDofHolder(),
+      layout_->GetNumComponents());
   }
 
-  void SetDOFHolderData(Rank1View<const T, DeviceMemorySpace> values) override
+  void SetDOFHolderData(Rank2View<const T, DeviceMemorySpace> values) override
   {
     PCMS_ALWAYS_ASSERT(values.size() ==
                        static_cast<size_t>(layout_->OwnedSize()));
-    CopyDeviceRank1ViewToDeviceView(device_data_, values);
-    SyncBackend(make_const_array_view(device_data_));
+    CopyDeviceRank2ViewToDeviceView(device_data_, values);
+    SyncBackend(GetDOFHolderData());
   }
 
   std::shared_ptr<MeshFieldBackend<T>> GetMeshFieldBackend() const
@@ -69,21 +73,25 @@ class MeshFieldsFieldData : public FieldData<T>
   }
 
 private:
-  void SyncBackend(Rank1View<const T, DeviceMemorySpace> flat)
+  void SyncBackend(Rank2View<const T, DeviceMemorySpace> data)
   {
     auto nodes_per_dim = layout_->GetNodesPerDim();
     auto num_components = layout_->GetNumComponents();
     auto& mesh = layout_->GetMesh();
-    size_t offset = 0;
+    // data is [dof_holder][component], contiguous node-major, so each mesh
+    // dimension owns a contiguous block of rows; SetData consumes a flat
+    // node-major span over that block.
+    size_t row_offset = 0;
     for (int i = 0; i <= mesh.dim(); ++i) {
       if (nodes_per_dim[i]) {
-        size_t len = static_cast<size_t>(mesh.nents(i)) *
-                     static_cast<size_t>(nodes_per_dim[i]) *
-                     static_cast<size_t>(num_components);
+        size_t num_rows = static_cast<size_t>(mesh.nents(i)) *
+                          static_cast<size_t>(nodes_per_dim[i]);
+        size_t len = num_rows * static_cast<size_t>(num_components);
         Rank1View<const T, DeviceMemorySpace> subspan{
-          flat.data_handle() + offset, len};
+          data.data_handle() + row_offset * static_cast<size_t>(num_components),
+          len};
         mesh_field_->SetData(subspan, nodes_per_dim[i], num_components, i);
-        offset += len;
+        row_offset += num_rows;
       }
     }
   }
diff --git a/src/pcms/field/data/point_cloud.cpp b/src/pcms/field/data/point_cloud.cpp
index 3e61355f..d333726b 100644
--- a/src/pcms/field/data/point_cloud.cpp
+++ b/src/pcms/field/data/point_cloud.cpp
@@ -10,10 +10,8 @@ namespace pcms
 PointCloud::PointCloud(std::shared_ptr<const PointCloudLayout> layout)
   : layout_(std::move(layout)),
     metadata_{},
-    device_data_("",
-                 layout_->GetDOFHolderCoordinates().GetCoordinates().extent(0)),
-    data_host_("",
-               layout_->GetDOFHolderCoordinates().GetCoordinates().extent(0))
+    device_data_("", layout_->GetDOFHolderCoordinates().GetValues().extent(0)),
+    data_host_("", layout_->GetDOFHolderCoordinates().GetValues().extent(0))
 {
 }
 
@@ -22,30 +20,34 @@ const FieldMetadata& PointCloud::GetMetadata() const
   return metadata_;
 }
 
-Rank1View<const Real, HostMemorySpace> PointCloud::GetDOFHolderDataHost() const
+Rank2View<const Real, HostMemorySpace> PointCloud::GetDOFHolderDataHost() const
 {
   Kokkos::deep_copy(data_host_, device_data_);
-  return make_const_array_view(data_host_);
+  const auto nc = layout_->GetNumComponents();
+  return Rank2View<const Real, HostMemorySpace>(
+    data_host_.data(), static_cast<LO>(data_host_.size()) / nc, nc);
 }
 
 void PointCloud::SetDOFHolderDataHost(
-  Rank1View<const Real, HostMemorySpace> data)
+  Rank2View<const Real, HostMemorySpace> data)
 {
   PCMS_FUNCTION_TIMER;
   PCMS_ALWAYS_ASSERT(data.size() == device_data_.size());
-  CopyHostRank1ViewToDeviceView(device_data_, data);
+  CopyHostRank2ViewToDeviceView(device_data_, data);
 }
 
-Rank1View<const Real, DeviceMemorySpace> PointCloud::GetDOFHolderData() const
+Rank2View<const Real, DeviceMemorySpace> PointCloud::GetDOFHolderData() const
 {
-  return make_const_array_view(device_data_);
+  const auto nc = layout_->GetNumComponents();
+  return Rank2View<const Real, DeviceMemorySpace>(
+    device_data_.data(), static_cast<LO>(device_data_.size()) / nc, nc);
 }
 
-void PointCloud::SetDOFHolderData(Rank1View<const Real, DeviceMemorySpace> data)
+void PointCloud::SetDOFHolderData(Rank2View<const Real, DeviceMemorySpace> data)
 {
   PCMS_FUNCTION_TIMER;
   PCMS_ALWAYS_ASSERT(data.size() == device_data_.size());
-  CopyDeviceRank1ViewToDeviceView(device_data_, data);
+  CopyDeviceRank2ViewToDeviceView(device_data_, data);
 }
 
 } // namespace pcms
diff --git a/src/pcms/field/data/point_cloud.h b/src/pcms/field/data/point_cloud.h
index dd0bc692..eedcc0b4 100644
--- a/src/pcms/field/data/point_cloud.h
+++ b/src/pcms/field/data/point_cloud.h
@@ -16,12 +16,12 @@ class PointCloud : public FieldData<Real>
 
   const FieldMetadata& GetMetadata() const override;
 
-  Rank1View<const Real, HostMemorySpace> GetDOFHolderDataHost() const override;
+  Rank2View<const Real, HostMemorySpace> GetDOFHolderDataHost() const override;
   void SetDOFHolderDataHost(
-    Rank1View<const Real, HostMemorySpace> data) override;
+    Rank2View<const Real, HostMemorySpace> data) override;
 
-  Rank1View<const Real, DeviceMemorySpace> GetDOFHolderData() const override;
-  void SetDOFHolderData(Rank1View<const Real, DeviceMemorySpace> data) override;
+  Rank2View<const Real, DeviceMemorySpace> GetDOFHolderData() const override;
+  void SetDOFHolderData(Rank2View<const Real, DeviceMemorySpace> data) override;
 
 private:
   std::shared_ptr<const PointCloudLayout> layout_;
diff --git a/src/pcms/field/data/simple.h b/src/pcms/field/data/simple.h
index cfefbd05..6deda15a 100644
--- a/src/pcms/field/data/simple.h
+++ b/src/pcms/field/data/simple.h
@@ -37,29 +37,33 @@ class SimpleFieldData : public FieldData<T>
 
   const FieldMetadata& GetMetadata() const override { return metadata_; }
 
-  Rank1View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
+  Rank2View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
   {
     Kokkos::deep_copy(host_data_, device_data_);
-    return make_const_array_view(host_data_);
+    return Rank2View<const T, HostMemorySpace>(host_data_.data(),
+                                               layout_->GetNumOwnedDofHolder(),
+                                               layout_->GetNumComponents());
   }
 
-  void SetDOFHolderDataHost(Rank1View<const T, HostMemorySpace> values) override
+  void SetDOFHolderDataHost(Rank2View<const T, HostMemorySpace> values) override
   {
     PCMS_ALWAYS_ASSERT(values.size() ==
                        static_cast<size_t>(layout_->OwnedSize()));
-    CopyHostRank1ViewToDeviceView(device_data_, values);
+    CopyHostRank2ViewToDeviceView(device_data_, values);
   }
 
-  Rank1View<const T, DeviceMemorySpace> GetDOFHolderData() const override
+  Rank2View<const T, DeviceMemorySpace> GetDOFHolderData() const override
   {
-    return make_const_array_view(device_data_);
+    return Rank2View<const T, DeviceMemorySpace>(
+      device_data_.data(), layout_->GetNumOwnedDofHolder(),
+      layout_->GetNumComponents());
   }
 
-  void SetDOFHolderData(Rank1View<const T, DeviceMemorySpace> values) override
+  void SetDOFHolderData(Rank2View<const T, DeviceMemorySpace> values) override
   {
     PCMS_ALWAYS_ASSERT(values.size() ==
                        static_cast<size_t>(layout_->OwnedSize()));
-    CopyDeviceRank1ViewToDeviceView(device_data_, values);
+    CopyDeviceRank2ViewToDeviceView(device_data_, values);
   }
 
 private:
diff --git a/src/pcms/field/data/xgc.h b/src/pcms/field/data/xgc.h
index 364b6042..cb7281a2 100644
--- a/src/pcms/field/data/xgc.h
+++ b/src/pcms/field/data/xgc.h
@@ -40,23 +40,28 @@ class XGCFieldData : public FieldData<T>
 
   const FieldMetadata& GetMetadata() const override { return metadata_; }
 
-  Rank1View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
+  Rank2View<const T, HostMemorySpace> GetDOFHolderDataHost() const override
   {
-    return Rank1View<const T, HostMemorySpace>(data_.data_handle(),
-                                               data_.size());
+    const auto nc = layout_->GetNumComponents();
+    return Rank2View<const T, HostMemorySpace>(
+      data_.data_handle(), static_cast<LO>(data_.size()) / nc, nc);
   }
 
-  void SetDOFHolderDataHost(Rank1View<const T, HostMemorySpace> values) override
+  void SetDOFHolderDataHost(Rank2View<const T, HostMemorySpace> values) override
   {
     if (values.size() != data_.size()) {
       throw pcms_error("XGCFieldData::SetDOFHolderDataHost: size mismatch");
     }
-    for (size_t i = 0; i < values.size(); ++i) {
-      data_(i) = values[i];
+    const auto num_dof = values.extent(0);
+    const auto num_comp = values.extent(1);
+    for (size_t i = 0; i < num_dof; ++i) {
+      for (size_t c = 0; c < num_comp; ++c) {
+        data_(i * num_comp + c) = values(i, c);
+      }
     }
   }
 
-  Rank1View<const T, DeviceMemorySpace> GetDOFHolderData() const override
+  Rank2View<const T, DeviceMemorySpace> GetDOFHolderData() const override
   {
     Kokkos::View<T*, HostMemorySpace> host_view("GetDOFHolderData_host_view",
                                                 data_.size());
@@ -64,13 +69,17 @@ class XGCFieldData : public FieldData<T>
       host_view(i) = data_(i);
     }
     Kokkos::deep_copy(device_data_, host_view);
-    return make_const_array_view(device_data_);
+    const auto nc = layout_->GetNumComponents();
+    return Rank2View<const T, DeviceMemorySpace>(
+      device_data_.data(), static_cast<LO>(device_data_.extent(0)) / nc, nc);
   }
 
-  void SetDOFHolderData(Rank1View<const T, DeviceMemorySpace> values) override
+  void SetDOFHolderData(Rank2View<const T, DeviceMemorySpace> values) override
   {
-    CopyDeviceRank1ViewToDeviceView(device_data_, values);
-    CopyRank1ViewToHost(data_, values);
+    // Fill the node-major device scratch from the (possibly non-node-major)
+    // 2D view, then mirror it back into the canonical host storage.
+    CopyDeviceRank2ViewToDeviceView(device_data_, values);
+    CopyRank1ViewToHost(data_, make_const_array_view(device_data_));
   }
 
 private:
diff --git a/src/pcms/field/eqdsk_field.h b/src/pcms/field/eqdsk_field.h
index 55c43298..bb4188c8 100644
--- a/src/pcms/field/eqdsk_field.h
+++ b/src/pcms/field/eqdsk_field.h
@@ -44,8 +44,8 @@ inline EQDSKField MakeEQDSKField(const EQDSKData& eqdsk_data)
   auto space = SplineFunctionSpace::FromUniformGrid(
     eqdsk_data.grid, CoordinateSystem::Cartesian);
   auto field = space.CreateField<Real>();
-  field.GetData().SetDOFHolderData(Rank1View<const Real, DeviceMemorySpace>(
-    eqdsk_data.PSIZR.data(), eqdsk_data.PSIZR.extent(0)));
+  field.GetData().SetDOFHolderData(Rank2View<const Real, DeviceMemorySpace>(
+    eqdsk_data.PSIZR.data(), eqdsk_data.PSIZR.extent(0), 1));
   return {std::move(space), std::move(field)};
 }
 
diff --git a/src/pcms/field/evaluator/mesh_fields.h b/src/pcms/field/evaluator/mesh_fields.h
index cbe69f5b..adf11c5b 100644
--- a/src/pcms/field/evaluator/mesh_fields.h
+++ b/src/pcms/field/evaluator/mesh_fields.h
@@ -128,7 +128,7 @@ class MeshFieldsEvaluatorFactory : public FieldEvaluatorFactory<T>
         "MeshFieldsEvaluatorFactory: NearestBoundary is not supported");
     }
 
-    auto coordinates = coords.GetCoordinates();
+    auto coordinates = coords.GetValues();
     Kokkos::View<Real* [2], DeviceMemorySpace> coords_search(
       "coords_search", coordinates.extent(0));
     Kokkos::parallel_for(
diff --git a/src/pcms/field/evaluator/mls_point_cloud.h b/src/pcms/field/evaluator/mls_point_cloud.h
index 3feeee9e..7000bd00 100644
--- a/src/pcms/field/evaluator/mls_point_cloud.h
+++ b/src/pcms/field/evaluator/mls_point_cloud.h
@@ -54,7 +54,7 @@ class MLSPointEvaluator : public PointEvaluator<Real, LayoutPolicy>
     Kokkos::parallel_for(
       "CopyDeviceDataToOmegaHWrite",
       Kokkos::RangePolicy<DeviceMemorySpace::execution_space>(0, n_sources),
-      KOKKOS_LAMBDA(int i) { src_w[i] = device_data(i); });
+      KOKKOS_LAMBDA(int i) { src_w[i] = device_data(i, 0); });
     Omega_h::Reals source_values(src_w);
 
     auto result = mls_interpolation(
diff --git a/src/pcms/field/evaluator/omega_h_lagrange.h b/src/pcms/field/evaluator/omega_h_lagrange.h
index 827a99d3..fd53d566 100644
--- a/src/pcms/field/evaluator/omega_h_lagrange.h
+++ b/src/pcms/field/evaluator/omega_h_lagrange.h
@@ -194,7 +194,7 @@ class OmegaHLagrangePointEvaluator : public PointEvaluator<T, LayoutPolicy>
           LO orig = hint_.orig_indices(k);
           LO elem = hint_.elem_ids(k);
           for (int c = 0; c < n_comp; ++c) {
-            values(orig, c) = dof_data[elem * n_comp + c];
+            values(orig, c) = dof_data(elem, c);
           }
         });
 
@@ -215,8 +215,7 @@ class OmegaHLagrangePointEvaluator : public PointEvaluator<T, LayoutPolicy>
             T val = T{};
             for (int v = 0; v < nvpe; ++v) {
               LO vert = elem_verts[elem * nvpe + v];
-              val +=
-                static_cast<T>(hint_.bary(k, v)) * dof_data[vert * n_comp + c];
+              val += static_cast<T>(hint_.bary(k, v)) * dof_data(vert, c);
             }
             values(orig, c) = val;
           }
@@ -288,7 +287,7 @@ class OmegaHLagrangeEvaluatorFactory : public FieldEvaluatorFactory<T>
         "OmegaHLagrangeEvaluatorFactory: NearestBoundary is not supported");
     }
 
-    auto raw_coords = coords.GetCoordinates();
+    auto raw_coords = coords.GetValues();
     LO n_pts = static_cast<LO>(raw_coords.extent(0));
     int mesh_dim = layout_->GetMesh().dim();
 
diff --git a/src/pcms/field/evaluator/point_cloud.h b/src/pcms/field/evaluator/point_cloud.h
index aafb8881..2ea37d51 100644
--- a/src/pcms/field/evaluator/point_cloud.h
+++ b/src/pcms/field/evaluator/point_cloud.h
@@ -72,13 +72,13 @@ class PointCloudEvaluatorFactory : public FieldEvaluatorFactory<Real>
     }
 
     // Extract source coordinates for the MLS solve.
-    const auto src_view = layout_->GetDOFHolderCoordinates().GetCoordinates();
+    const auto src_view = layout_->GetDOFHolderCoordinates().GetValues();
     const int dim = layout_->GetDimension();
     Omega_h::Reals source_coords =
       flatten_to_omega_h_reals(src_view, "src_coords");
 
     // Extract target coordinates for the MLS solve.
-    const auto tgt_view = coords.GetCoordinates();
+    const auto tgt_view = coords.GetValues();
     PCMS_ALWAYS_ASSERT(static_cast<int>(tgt_view.extent(1)) == dim);
     Omega_h::Reals target_coords_oh =
       flatten_to_omega_h_reals(tgt_view, "tgt_coords");
diff --git a/src/pcms/field/evaluator/uniform_grid.h b/src/pcms/field/evaluator/uniform_grid.h
index 38299162..ed061458 100644
--- a/src/pcms/field/evaluator/uniform_grid.h
+++ b/src/pcms/field/evaluator/uniform_grid.h
@@ -99,7 +99,7 @@ class UniformGridPointEvaluator : public PointEvaluator<Real, LayoutPolicy>
             values(i, c) = fv;
           } else {
             LO dof_idx = cell_indices(i);
-            values(i, c) = dof_data[dof_idx * n_comp + c];
+            values(i, c) = dof_data(dof_idx, c);
           }
         });
       return;
@@ -164,7 +164,7 @@ class UniformGridPointEvaluator : public PointEvaluator<Real, LayoutPolicy>
       "copy_dof_data_to_values_interp",
       Kokkos::RangePolicy<DeviceMemorySpace::execution_space>(
         0, static_cast<LO>(dof_data.size())),
-      KOKKOS_LAMBDA(const LO i) { values_interp(i) = dof_data[i]; });
+      KOKKOS_LAMBDA(const LO i) { values_interp(i) = dof_data(i, 0); });
 
     auto interpolator = RegularGridInterpolator(
       parametric_coords, values_interp, cell_indices_interp, dimensions_view);
@@ -186,7 +186,7 @@ class UniformGridPointEvaluator : public PointEvaluator<Real, LayoutPolicy>
 
 private:
   std::shared_ptr<const UniformGridFieldLayout<Dim>> layout_;
-  const UniformGrid<Dim>& grid_;
+  UniformGrid<Dim> grid_;
   UniformGridFieldLocalizationHint<Dim> hint_;
   Real fill_value_;
 };
@@ -241,7 +241,7 @@ class UniformGridEvaluatorFactory : public FieldEvaluatorFactory<Real>
         "not implemented");
     }
 
-    auto coordinates = coords.GetCoordinates();
+    auto coordinates = coords.GetValues();
     LO num_points = static_cast<LO>(coordinates.extent(0));
 
     Kokkos::View<LO*, DeviceMemorySpace> cell_indices_device("cell_indices",
@@ -304,7 +304,7 @@ class UniformGridEvaluatorFactory : public FieldEvaluatorFactory<Real>
 
 private:
   std::shared_ptr<const UniformGridFieldLayout<Dim>> layout_;
-  const UniformGrid<Dim>& grid_;
+  UniformGrid<Dim> grid_;
 };
 
 using UniformGridEvaluatorFactory2D = UniformGridEvaluatorFactory<2>;
diff --git a/src/pcms/field/evaluator/uniform_grid_spline.h b/src/pcms/field/evaluator/uniform_grid_spline.h
index 576f000a..3fb693ad 100644
--- a/src/pcms/field/evaluator/uniform_grid_spline.h
+++ b/src/pcms/field/evaluator/uniform_grid_spline.h
@@ -86,7 +86,7 @@ class UniformGridSplinePointEvaluator2D
       "copy_spline_values_device",
       Kokkos::MDRangePolicy<Kokkos::Rank<2>>({0, 0}, {ny, nx}),
       KOKKOS_LAMBDA(const LO iy, const LO ix) {
-        spline_values(ix * ny + iy) = dof_data[iy * nx + ix];
+        spline_values(ix * ny + iy) = dof_data(iy * nx + ix, 0);
       });
 
     if (hint_.num_out_of_bounds_ == static_cast<size_t>(num_points) &&
@@ -171,7 +171,7 @@ class UniformGridSplinePointEvaluator2D
 
 private:
   std::shared_ptr<const UniformGridFieldLayout<2>> layout_;
-  const UniformGrid<2>& grid_;
+  UniformGrid<2> grid_;
   UniformGridFieldLocalizationHint<2> hint_;
   Real fill_value_;
   Kokkos::View<Real*, DeviceMemorySpace> x_coords_;
@@ -223,7 +223,7 @@ class UniformGridSplineEvaluatorFactory2D : public FieldEvaluatorFactory<Real>
         "single-component fields");
     }
 
-    auto coordinates = coords.GetCoordinates();
+    auto coordinates = coords.GetValues();
     LO num_points = static_cast<LO>(coordinates.extent(0));
 
     Kokkos::View<Real**, DeviceMemorySpace> coordinates_d("coordinates_d",
@@ -287,7 +287,7 @@ class UniformGridSplineEvaluatorFactory2D : public FieldEvaluatorFactory<Real>
 
 private:
   std::shared_ptr<const UniformGridFieldLayout<2>> layout_;
-  const UniformGrid<2>& grid_;
+  UniformGrid<2> grid_;
 };
 
 } // namespace pcms
diff --git a/src/pcms/field/field.h b/src/pcms/field/field.h
index 98f0301f..433717d0 100644
--- a/src/pcms/field/field.h
+++ b/src/pcms/field/field.h
@@ -35,22 +35,22 @@ class Field
 
   const FieldLayout& GetLayout() const { return *layout_; }
 
-  Rank1View<const T, HostMemorySpace> GetDOFHolderDataHost() const
+  Rank2View<const T, HostMemorySpace> GetDOFHolderDataHost() const
   {
     return data_->GetDOFHolderDataHost();
   }
 
-  void SetDOFHolderDataHost(Rank1View<const T, HostMemorySpace> v)
+  void SetDOFHolderDataHost(Rank2View<const T, HostMemorySpace> v)
   {
     data_->SetDOFHolderDataHost(v);
   }
 
-  Rank1View<const T, DeviceMemorySpace> GetDOFHolderData() const
+  Rank2View<const T, DeviceMemorySpace> GetDOFHolderData() const
   {
     return data_->GetDOFHolderData();
   }
 
-  void SetDOFHolderData(Rank1View<const T, DeviceMemorySpace> v)
+  void SetDOFHolderData(Rank2View<const T, DeviceMemorySpace> v)
   {
     data_->SetDOFHolderData(v);
   }
diff --git a/src/pcms/field/field_data.h b/src/pcms/field/field_data.h
index eee02a00..fb3846cd 100644
--- a/src/pcms/field/field_data.h
+++ b/src/pcms/field/field_data.h
@@ -30,7 +30,11 @@ namespace pcms
 //
 // Contract:
 //   - GetMetadata() describes the stored coefficient data.
-//   - The flattened DOF-holder data ordering must match the layout ordering.
+//   - DOF-holder data is exposed as a 2D [num_dof_holder][num_components] view
+//     (the canonical shape); the DOF-holder ordering must match the layout.
+//     The default layout_right means it is contiguous node-major (component
+//     fastest, i.e. flat index i*num_components + c) on every backend, so a
+//     flat/wire representation is just the contiguous data_handle().
 //   - Views returned by GetDOFHolderDataHost()/GetDOFHolderData() remain
 //     valid until this FieldData object is mutated or destroyed.
 //   - SetDOFHolderDataHost()/SetDOFHolderData() replace the entire stored
@@ -50,18 +54,18 @@ class FieldData
 
   virtual const FieldMetadata& GetMetadata() const = 0;
 
-  // Direct access to flattened DOFHolder-ordered coefficient data.
-  // The returned view remains valid until the FieldData is mutated or
-  // destroyed.
-  virtual Rank1View<const T, HostMemorySpace> GetDOFHolderDataHost() const = 0;
+  // Direct access to DOF-holder coefficient data as a 2D
+  // [num_dof_holder][num_components] view. The returned view remains valid
+  // until the FieldData is mutated or destroyed.
+  virtual Rank2View<const T, HostMemorySpace> GetDOFHolderDataHost() const = 0;
   virtual void SetDOFHolderDataHost(
-    Rank1View<const T, HostMemorySpace> values) = 0;
+    Rank2View<const T, HostMemorySpace> values) = 0;
 
   // The returned view remains valid until the FieldData is mutated or
   // destroyed.
-  virtual Rank1View<const T, DeviceMemorySpace> GetDOFHolderData() const = 0;
+  virtual Rank2View<const T, DeviceMemorySpace> GetDOFHolderData() const = 0;
   virtual void SetDOFHolderData(
-    Rank1View<const T, DeviceMemorySpace> values) = 0;
+    Rank2View<const T, DeviceMemorySpace> values) = 0;
 
   virtual ~FieldData() noexcept = default;
 };
diff --git a/src/pcms/field/layout/omega_h_lagrange.cpp b/src/pcms/field/layout/omega_h_lagrange.cpp
index 93ebd0bd..a2749a6e 100644
--- a/src/pcms/field/layout/omega_h_lagrange.cpp
+++ b/src/pcms/field/layout/omega_h_lagrange.cpp
@@ -3,6 +3,7 @@
 #include "pcms/utility/mesh_geometry.h"
 #include "pcms/utility/omega_h_array_utils.h"
 #include "pcms/utility/profile.h"
+#include <Omega_h_class.hpp>
 #include <stdexcept>
 
 namespace pcms
@@ -48,6 +49,35 @@ Omega_h::Write<Omega_h::GO> BuildGids(Omega_h::Mesh& mesh, int entity_dim,
   return gids;
 }
 
+// Some meshes (e.g. certain XGC meshes) are stored without any geometric
+// classification (no class_id/class_dim tags). The layout and its
+// discretization read these tags on every mesh dimension, so derive a default
+// classification when it is missing. classify_elements +
+// finalize_classification populate a geometrically meaningful class_dim on all
+// dimensions but do not create class_id (that normally comes from a geometric
+// model we don't have here), so fill any missing class_id with 0.
+// Classification is only consulted for coupling DOF matching
+// (field_exchange_planner); standalone field transfer never reads it, so a
+// default id is sufficient to keep such meshes usable.
+void EnsureClassification(Omega_h::Mesh& mesh)
+{
+  if (mesh.has_tag(mesh.dim(), "class_id") &&
+      mesh.has_tag(mesh.dim(), "class_dim")) {
+    return;
+  }
+  if (!mesh.has_tag(mesh.dim(), "class_dim")) {
+    Omega_h::classify_elements(&mesh);
+    Omega_h::finalize_classification(&mesh);
+  }
+  for (int d = 0; d <= mesh.dim(); ++d) {
+    if (!mesh.has_tag(d, "class_id")) {
+      mesh.add_tag<Omega_h::ClassId>(
+        d, "class_id", 1,
+        Omega_h::Read<Omega_h::ClassId>(mesh.nents(d), 0, "class_id"));
+    }
+  }
+}
+
 Kokkos::View<bool*, DeviceMemorySpace> BuildOwned(Omega_h::Mesh& mesh,
                                                   int entity_dim)
 {
@@ -87,6 +117,7 @@ OmegaHLagrangeLayout::OmegaHLagrangeLayout(Omega_h::Mesh& mesh, int order,
   PCMS_FUNCTION_TIMER;
   int entity_dim = EntityDimForOrder(order_, mesh_.dim());
 
+  EnsureClassification(mesh_);
   gids_ = BuildGids(mesh_, entity_dim, global_id_name_);
   gids_host_ = Omega_h::HostWrite<Omega_h::GO>(gids_);
   coords_ = get_entity_centroids(mesh_, entity_dim);
@@ -130,6 +161,7 @@ OmegaHLagrangeLayout::OmegaHLagrangeLayout(
   PCMS_FUNCTION_TIMER;
   int entity_dim = EntityDimForOrder(order_, mesh_.dim());
 
+  EnsureClassification(mesh_);
   gids_ = BuildGids(mesh_, entity_dim, global_id_name_);
   gids_host_ = Omega_h::HostWrite<Omega_h::GO>(gids_);
   coords_ = get_entity_centroids(mesh_, entity_dim);
diff --git a/src/pcms/field/uniform_grid_binary_field.h b/src/pcms/field/uniform_grid_binary_field.h
index 4bdd7d27..6a71de7a 100644
--- a/src/pcms/field/uniform_grid_binary_field.h
+++ b/src/pcms/field/uniform_grid_binary_field.h
@@ -44,7 +44,7 @@ CreateUniformGridBinaryField(Omega_h::Mesh& mesh, const UniformGrid<Dim>& grid)
   auto field = function_space.template CreateField<Real>(FieldMetadata{});
 
   auto coord_view = layout->GetDOFHolderCoordinates();
-  auto coords = coord_view.GetCoordinates();
+  auto coords = coord_view.GetValues();
   LO n = layout->GetNumOwnedDofHolder();
 
   Kokkos::View<Real* [Dim]> coords_d("coords_d", n);
@@ -73,7 +73,7 @@ CreateUniformGridBinaryField(Omega_h::Mesh& mesh, const UniformGrid<Dim>& grid)
     data(i) = (results_h(i).element_id >= 0) ? 1.0 : 0.0;
 
   field.SetDOFHolderDataHost(
-    Rank1View<const Real, HostMemorySpace>(data.data(), n));
+    Rank2View<const Real, HostMemorySpace>(data.data(), n, 1));
 
   return {std::move(layout), std::move(field)};
 }
diff --git a/src/pcms/localization/adj_search.cpp b/src/pcms/localization/adj_search.cpp
index e1b937d7..6f88a2ce 100644
--- a/src/pcms/localization/adj_search.cpp
+++ b/src/pcms/localization/adj_search.cpp
@@ -160,7 +160,10 @@ OMEGA_H_INLINE void add_support_if_unvisited_and_within_cutoff(
   if (!visited.notVisited(candidate_id))
     return;
 
-  visited.push_back(candidate_id);
+  // If the visited buffer is full, stop here so the BFS terminates. Without
+  // this, unrecorded vertices are treated as unvisited and re-queued forever.
+  if (!visited.push_back(candidate_id))
+    return;
 
   Omega_h::Real candidate_coords[max_dim];
   for (Omega_h::LO k = 0; k < dim; ++k) {
diff --git a/src/pcms/localization/localization_path_selection.h b/src/pcms/localization/localization_path_selection.h
index 40bb9894..e08069cb 100644
--- a/src/pcms/localization/localization_path_selection.h
+++ b/src/pcms/localization/localization_path_selection.h
@@ -44,8 +44,8 @@ inline std::optional<int> GetUniformDofHolderEntityDim(
     return std::nullopt;
   }
 
-  if (static_cast<LO>(layout.GetDOFHolderCoordinates().GetCoordinates().extent(
-        0)) != disc->GetNumEntities(*entity_dim)) {
+  if (static_cast<LO>(layout.GetDOFHolderCoordinates().GetValues().extent(0)) !=
+      disc->GetNumEntities(*entity_dim)) {
     return std::nullopt;
   }
 
diff --git a/src/pcms/localization/mesh_localization.cpp b/src/pcms/localization/mesh_localization.cpp
index a73d21bd..27821430 100644
--- a/src/pcms/localization/mesh_localization.cpp
+++ b/src/pcms/localization/mesh_localization.cpp
@@ -20,7 +20,7 @@ SupportResults AdjacencyLocalizationFactory::Build(
       "AdjacencyLocalizationFactory: only Cartesian coordinates are supported");
   }
 
-  const auto tgt_view = target_coords.GetCoordinates();
+  const auto tgt_view = target_coords.GetValues();
   const int dim = source_mesh_.dim();
   PCMS_ALWAYS_ASSERT(static_cast<int>(tgt_view.extent(1)) == dim);
   const int n_tgt = static_cast<int>(tgt_view.extent(0));
diff --git a/src/pcms/localization/point_cloud_localization.h b/src/pcms/localization/point_cloud_localization.h
index 5a1d6e81..401b17e7 100644
--- a/src/pcms/localization/point_cloud_localization.h
+++ b/src/pcms/localization/point_cloud_localization.h
@@ -41,7 +41,7 @@ class PointCloudLocalizationFactory : public LocalizationFactory
     Omega_h::Reals source_coords =
       flatten_to_omega_h_reals(src_view, "src_coords");
 
-    const auto tgt_view = target_coords.GetCoordinates();
+    const auto tgt_view = target_coords.GetValues();
     PCMS_ALWAYS_ASSERT(static_cast<int>(tgt_view.extent(1)) == dim);
     Omega_h::Reals target_coords_oh =
       flatten_to_omega_h_reals(tgt_view, "tgt_coords");
diff --git a/src/pcms/localization/queue_visited.hpp b/src/pcms/localization/queue_visited.hpp
index 421281c5..8d768eec 100644
--- a/src/pcms/localization/queue_visited.hpp
+++ b/src/pcms/localization/queue_visited.hpp
@@ -56,7 +56,7 @@ class Track
   ~Track() {}
 
   OMEGA_H_INLINE
-  void push_back(const int& item);
+  bool push_back(const int& item);
 
   OMEGA_H_INLINE
   int size();
@@ -107,15 +107,18 @@ bool Queue::isFull() const
 }
 
 OMEGA_H_INLINE
-void Track::push_back(const int& item)
+bool Track::push_back(const int& item)
 {
   if (count == MAX_SIZE_TRACK) {
-    printf("track is full %d\n", count);
-    return;
+    // Visited buffer is full. Report failure so callers can stop expanding the
+    // search; otherwise new vertices can never be marked visited and the BFS
+    // re-queues them forever (infinite loop).
+    return false;
   }
   last = (last + 1) % MAX_SIZE_TRACK;
   tracking_array[last] = item;
   count++;
+  return true;
 }
 
 OMEGA_H_INLINE
diff --git a/src/pcms/pythonapi/bind_field_base.cpp b/src/pcms/pythonapi/bind_field_base.cpp
index 4854ca80..ee4bff09 100644
--- a/src/pcms/pythonapi/bind_field_base.cpp
+++ b/src/pcms/pythonapi/bind_field_base.cpp
@@ -76,7 +76,7 @@ void bind_coordinate_system_module(py::module& m)
     .def(
       "get_coordinates",
       [](const CoordinateView<HostMemorySpace>& self) {
-        auto coords = self.GetCoordinates();
+        auto coords = self.GetValues();
         // Convert to numpy array
         py::array_t<Real> result({static_cast<py::ssize_t>(coords.extent(0)),
                                   static_cast<py::ssize_t>(coords.extent(1))});
@@ -207,11 +207,16 @@ void bind_create_field_module(py::module& m)
       "get_dof_holder_data",
       [](const Field<Real>& self) {
         auto data = self.GetDOFHolderDataHost();
+        // Flatten the [dof][comp] data into node-major order for the 1D numpy
+        // array (Python-facing DOF data stays 1D).
+        const auto num_dof = data.extent(0);
+        const auto num_comp = data.extent(1);
         py::array_t<Real> result(static_cast<py::ssize_t>(data.size()));
         auto buf = result.request();
         Real* ptr = static_cast<Real*>(buf.ptr);
-        for (size_t i = 0; i < data.size(); ++i)
-          ptr[i] = data[i];
+        for (size_t i = 0; i < num_dof; ++i)
+          for (size_t c = 0; c < num_comp; ++c)
+            ptr[i * num_comp + c] = data(i, c);
         return result;
       },
       "Get the DOF holder data as a 1D numpy array")
@@ -223,9 +228,11 @@ void bind_create_field_module(py::module& m)
         if (buf.ndim != 1) {
           throw std::runtime_error("DOF holder data must be a 1D array");
         }
-        Rank1View<const Real, HostMemorySpace> view(
-          static_cast<const Real*>(buf.ptr), static_cast<size_t>(buf.shape[0]));
-        self.SetDOFHolderDataHost(view);
+        // Reshape the flat 1D input into the field's [dof][comp] layout.
+        const int nc = self.GetLayout().GetNumComponents();
+        const auto total = static_cast<size_t>(buf.shape[0]);
+        self.SetDOFHolderDataHost(Rank2View<const Real, HostMemorySpace>(
+          static_cast<const Real*>(buf.ptr), static_cast<LO>(total / nc), nc));
       },
       py::arg("data"), "Set the DOF holder data from a 1D numpy array")
 
@@ -247,7 +254,7 @@ void bind_create_field_module(py::module& m)
       "get_dof_holder_coordinates",
       [](const Field<Real>& self) {
         auto cv = self.GetLayout().GetDOFHolderCoordinates();
-        auto coords = cv.GetCoordinates();
+        auto coords = cv.GetValues();
         Kokkos::View<Real**, DeviceMemorySpace> coords_device(
           "coords_device", coords.extent(0), coords.extent(1));
         Kokkos::parallel_for(
diff --git a/src/pcms/transfer/interpolator.h b/src/pcms/transfer/interpolator.h
index d60b31bb..a8b5b07e 100644
--- a/src/pcms/transfer/interpolator.h
+++ b/src/pcms/transfer/interpolator.h
@@ -36,10 +36,9 @@ class Interpolator : public TransferOperator<T>
   // Expensive: localizes target DOF coords into source mesh. Called once.
   Interpolator(const FunctionSpace& source_space,
                const FunctionSpace& target_space, OutOfBoundsPolicy policy = {})
-    : num_points_(static_cast<LO>(target_space.GetLayout()
-                                    ->GetDOFHolderCoordinates()
-                                    .GetCoordinates()
-                                    .extent(0))),
+    : num_points_(static_cast<LO>(
+        target_space.GetLayout()->GetDOFHolderCoordinates().GetValues().extent(
+          0))),
       n_comp_(target_space.GetLayout()->GetNumComponents()),
       evaluator_(source_space.CreatePointEvaluator<T>(
         EvaluationRequest::FromFunctionSpace(target_space, policy)))
@@ -66,7 +65,8 @@ class Interpolator : public TransferOperator<T>
           flat(i * n_comp + c) = output(i, c);
         }
       });
-    target.GetData().SetDOFHolderData(make_const_array_view(flat));
+    target.GetData().SetDOFHolderData(
+      Rank2View<const T, DeviceMemorySpace>(flat.data(), num_points, n_comp));
   }
 
 private:
diff --git a/src/pcms/transfer/omega_h_conservative_projection.cpp b/src/pcms/transfer/omega_h_conservative_projection.cpp
index 3b70e427..c7ae7e65 100644
--- a/src/pcms/transfer/omega_h_conservative_projection.cpp
+++ b/src/pcms/transfer/omega_h_conservative_projection.cpp
@@ -88,12 +88,15 @@ void OmegaHConservativeProjection::Apply(const Field<Real>& source,
 {
   CheckApplyCompatible(source, target, *source_layout_, *target_layout_);
 
-  const auto source_values = MakeOmegaHReals(source.GetDOFHolderDataHost());
+  const auto source_values =
+    MakeOmegaHReals(FlattenToRank1View(source.GetDOFHolderDataHost()));
   const auto target_values = solveGalerkinProjection(
     target_layout_->GetMesh(), source_layout_->GetMesh(), intersections_,
     source_values);
   auto target_values_h = Omega_h::HostRead<Omega_h::Real>(target_values);
-  target.SetDOFHolderDataHost(make_const_array_view(target_values_h));
+  target.SetDOFHolderDataHost(Rank2View<const Real, HostMemorySpace>(
+    target_values_h.data(), target_layout_->GetNumOwnedDofHolder(),
+    target_layout_->GetNumComponents()));
 }
 
 } // namespace pcms
diff --git a/src/pcms/utility/arrays.h b/src/pcms/utility/arrays.h
index 29e0d426..96a5828d 100644
--- a/src/pcms/utility/arrays.h
+++ b/src/pcms/utility/arrays.h
@@ -243,6 +243,18 @@ auto MakeConstRank2View(Omega_h::Read<T> array, int dim)
     array.data(), array.size() / dim, dim);
 }
 
+// Flatten a contiguous (exhaustive) 2D [dof][comp] view into a 1D view over the
+// same storage. Intended for boundary code (serialization, element-wise
+// comparison) that wants a flat [num_dof_holder * num_components] view rather
+// than indexing pointers directly.
+template <typename ElementType, typename MemorySpace, typename LayoutPolicy>
+Rank1View<ElementType, MemorySpace> FlattenToRank1View(
+  Rank2View<ElementType, MemorySpace, LayoutPolicy> view)
+{
+  return Rank1View<ElementType, MemorySpace>(view.data_handle(),
+                                             static_cast<LO>(view.size()));
+}
+
 // utility function to deep copy between layout incompatible views
 // layout imcompatible view can't be deep_copied directly between different
 // memory spaces, the workaround is provided from
@@ -338,6 +350,85 @@ void CopyHostRank1ViewToDeviceView(Kokkos::View<T*, DeviceMemorySpace> dest,
   Kokkos::deep_copy(dest, src_tmp);
 }
 
+// Copy a [dof][comp] device Rank2View into a flat node-major device View.
+// Uses the 2D indexing src(i, c) so the result is correct regardless of the
+// source view's layout (unlike flattening through data_handle(), which assumes
+// node-major storage and silently reorders wide/multi-component data on GPU).
+template <typename T>
+void CopyDeviceRank2ViewToDeviceView(Kokkos::View<T*, DeviceMemorySpace> dest,
+                                     Rank2View<const T, DeviceMemorySpace> src)
+{
+  const size_t num_dof = src.extent(0);
+  const size_t num_comp = src.extent(1);
+  if (dest.extent(0) != num_dof * num_comp) {
+    throw pcms_error("CopyDeviceRank2ViewToDeviceView: size mismatch");
+  }
+  Kokkos::parallel_for(
+    "CopyDeviceRank2ViewToDeviceView", num_dof, KOKKOS_LAMBDA(LO i) {
+      for (size_t c = 0; c < num_comp; ++c) {
+        dest(i * num_comp + c) = src(i, c);
+      }
+    });
+}
+
+// Rank-2 destination overload, so callers can migrate their internal storage
+// from a flat View<T*> to a View<T**> without changing the call site.
+template <typename T>
+void CopyDeviceRank2ViewToDeviceView(Kokkos::View<T**, DeviceMemorySpace> dest,
+                                     Rank2View<const T, DeviceMemorySpace> src)
+{
+  const size_t num_dof = src.extent(0);
+  const size_t num_comp = src.extent(1);
+  if (dest.extent(0) != num_dof || dest.extent(1) != num_comp) {
+    throw pcms_error("CopyDeviceRank2ViewToDeviceView: size mismatch");
+  }
+  Kokkos::parallel_for(
+    "CopyDeviceRank2ViewToDeviceView", num_dof, KOKKOS_LAMBDA(LO i) {
+      for (size_t c = 0; c < num_comp; ++c) {
+        dest(i, c) = src(i, c);
+      }
+    });
+}
+
+// Host analogue: copy a [dof][comp] host Rank2View into a flat node-major
+// device View. The host Rank2View is layout_right (contiguous node-major), so
+// we alias its storage with an unmanaged view and deep_copy straight to device.
+template <typename T>
+void CopyHostRank2ViewToDeviceView(Kokkos::View<T*, DeviceMemorySpace> dest,
+                                   Rank2View<const T, HostMemorySpace> src)
+{
+  const size_t num_dof = src.extent(0);
+  const size_t num_comp = src.extent(1);
+  if (dest.extent(0) != num_dof * num_comp) {
+    throw pcms_error("CopyHostRank2ViewToDeviceView: size mismatch");
+  }
+  Kokkos::View<const T*, HostMemorySpace,
+               Kokkos::MemoryTraits<Kokkos::Unmanaged>>
+    src_view(src.data_handle(), num_dof * num_comp);
+  Kokkos::deep_copy(dest, src_view);
+}
+
+// Rank-2 destination overload for host-to-device copies. The device View<T**>
+// may use a different layout than the host source, so stage through a mirror
+// that matches the destination layout.
+template <typename T>
+void CopyHostRank2ViewToDeviceView(Kokkos::View<T**, DeviceMemorySpace> dest,
+                                   Rank2View<const T, HostMemorySpace> src)
+{
+  const size_t num_dof = src.extent(0);
+  const size_t num_comp = src.extent(1);
+  if (dest.extent(0) != num_dof || dest.extent(1) != num_comp) {
+    throw pcms_error("CopyHostRank2ViewToDeviceView: size mismatch");
+  }
+  auto src_mirror = Kokkos::create_mirror_view(dest);
+  for (size_t i = 0; i < num_dof; ++i) {
+    for (size_t c = 0; c < num_comp; ++c) {
+      src_mirror(i, c) = src(i, c);
+    }
+  }
+  Kokkos::deep_copy(dest, src_mirror);
+}
+
 // utility function to copy from Rank1View to Rank1View
 template <typename T>
 void CopyRank1ViewToHost(Rank1View<T, HostMemorySpace> dest,
diff --git a/test/field_test_utils.h b/test/field_test_utils.h
index f3b2a9bd..3fbb624f 100644
--- a/test/field_test_utils.h
+++ b/test/field_test_utils.h
@@ -146,7 +146,7 @@ inline void SetField(const FieldLayout& layout, FieldData<Real>& field,
 {
   using MemorySpace = typename ExecutionSpace::memory_space;
 
-  auto dof_coords = layout.GetDOFHolderCoordinates().GetCoordinates();
+  auto dof_coords = layout.GetDOFHolderCoordinates().GetValues();
   int n = static_cast<int>(dof_coords.extent(0));
   Kokkos::View<Real* [2], MemorySpace> coords_device("coords_device", n);
   Kokkos::parallel_for(
@@ -164,7 +164,7 @@ inline void SetField(const FieldLayout& layout, FieldData<Real>& field,
   auto data_host =
     Kokkos::create_mirror_view_and_copy(HostMemorySpace(), data_device);
   field.SetDOFHolderDataHost(
-    Rank1View<const Real, HostMemorySpace>(data_host.data(), n));
+    Rank2View<const Real, HostMemorySpace>(data_host.data(), n, 1));
 }
 
 template <typename ExecutionSpace = DefaultExecutionSpace, typename Func>
@@ -185,7 +185,7 @@ inline void SetField(FieldData<Real>& field, const FieldLayout& layout,
 inline void CheckSerializeDeserialize(const FieldLayout& layout,
                                       FieldData<Real>& field)
 {
-  auto data_before = field.GetDOFHolderDataHost();
+  auto data_before = FlattenToRank1View(field.GetDOFHolderDataHost());
   int n = static_cast<int>(data_before.size());
 
   std::vector<Real> buffer(n);
@@ -201,7 +201,7 @@ inline void CheckSerializeDeserialize(const FieldLayout& layout,
   serializer.Deserialize(
     field, layout, Rank1View<const Real, HostMemorySpace>(buf_view), perm_view);
 
-  auto data_after = field.GetDOFHolderDataHost();
+  auto data_after = FlattenToRank1View(field.GetDOFHolderDataHost());
   REQUIRE(data_after.size() == data_before.size());
   for (int i = 0; i < n; ++i) {
     REQUIRE(data_after[i] == Catch::Approx(data_before[i]));
diff --git a/test/test_field_communication.cpp b/test/test_field_communication.cpp
index b27ef428..83b8bc5b 100644
--- a/test/test_field_communication.cpp
+++ b/test/test_field_communication.cpp
@@ -135,7 +135,8 @@ void client1(MPI_Comm comm, Omega_h::Mesh& mesh, std::string comm_name,
     [=](int i) { ids[i] = gids[i]; });
 
   auto field = factory.CreateField<Real>(pcms::FieldMetadata{});
-  field.SetDOFHolderDataHost(pcms::make_const_array_view(ids));
+  field.SetDOFHolderDataHost(
+    pcms::Rank2View<const Real, pcms::HostMemorySpace>(ids.data(), n, 1));
 
   pcms::FieldLayoutCommunicator layout_comm(comm_name + "1", comm, rdv, channel,
                                             *layout);
@@ -170,7 +171,7 @@ void client2(MPI_Comm comm, Omega_h::Mesh& mesh, std::string comm_name,
   field_comm.Receive();
   channel.EndReceiveCommunicationPhase();
 
-  auto copied_array = field.GetDOFHolderDataHost();
+  auto copied_array = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   auto owned = layout->GetOwnedHost();
 
   PCMS_ALWAYS_ASSERT(copied_array.size() == gids.size());
diff --git a/test/test_field_copy.cpp b/test/test_field_copy.cpp
index 988a62d4..9906bbbc 100644
--- a/test/test_field_copy.cpp
+++ b/test/test_field_copy.cpp
@@ -30,12 +30,14 @@ void test_copy(Omega_h::CommPtr world, int dim, int order, int num_components)
     [=](int i) { ids[i] = i; });
 
   auto original = factory.CreateField<Real>(pcms::FieldMetadata{});
-  original.SetDOFHolderDataHost(pcms::make_const_array_view(ids));
+  original.SetDOFHolderDataHost(
+    pcms::Rank2View<const Real, pcms::HostMemorySpace>(
+      ids.data(), layout->GetNumOwnedDofHolder(), num_components));
 
   auto copied = factory.CreateField<Real>(pcms::FieldMetadata{});
   pcms::Copy<Real> copy(factory, factory);
   copy.Apply(original, copied);
-  auto copied_array = copied.GetDOFHolderDataHost();
+  auto copied_array = pcms::FlattenToRank1View(copied.GetDOFHolderDataHost());
 
   REQUIRE(copied_array.size() == ndata);
   int sum = 0;
diff --git a/test/test_field_evaluation.cpp b/test/test_field_evaluation.cpp
index 2a2094a7..c4a29378 100644
--- a/test/test_field_evaluation.cpp
+++ b/test/test_field_evaluation.cpp
@@ -77,7 +77,8 @@ TEST_CASE("evaluate quadratic 2d meshfields_field")
   Omega_h::HostWrite<Real> test_f_host(test_f);
   auto field = factory.CreateField<Real>(pcms::FieldMetadata{});
   field.GetData().SetDOFHolderDataHost(
-    pcms::make_const_array_view(test_f_host));
+    pcms::Rank2View<const Real, pcms::HostMemorySpace>(test_f_host.data(),
+                                                       test_f_host.size(), 1));
 
   pcms::test::CheckEvaluation(
     factory, field, kEvalCoords,
diff --git a/test/test_field_interpolation.cpp b/test/test_field_interpolation.cpp
index dde9cf80..fdada727 100644
--- a/test/test_field_interpolation.cpp
+++ b/test/test_field_interpolation.cpp
@@ -34,8 +34,9 @@ TEST_CASE("interpolate linear 2d omega_h_field")
 
   pcms::Interpolator<Real> interp(factory, factory);
   interp.Apply(field, interpolated);
-  auto interpolated_dof = interpolated.GetDOFHolderDataHost();
-  auto original_dof = field.GetDOFHolderDataHost();
+  auto interpolated_dof =
+    pcms::FlattenToRank1View(interpolated.GetDOFHolderDataHost());
+  auto original_dof = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   REQUIRE(interpolated_dof.size() == original_dof.size());
   for (int i = 0; i < static_cast<int>(interpolated_dof.size()); ++i) {
     REQUIRE_THAT(interpolated_dof[i],
@@ -81,13 +82,15 @@ TEST_CASE("interpolate quadratic 2d meshfields_field")
   Omega_h::HostWrite<Real> test_f_host(test_f);
   auto field = factory2.CreateField<Real>(pcms::FieldMetadata{});
   auto interpolated = factory2.CreateField<Real>(pcms::FieldMetadata{});
-  field.SetDOFHolderDataHost(pcms::make_const_array_view(test_f_host));
+  field.SetDOFHolderDataHost(pcms::Rank2View<const Real, pcms::HostMemorySpace>(
+    test_f_host.data(), test_f_host.size(), 1));
 
   pcms::Interpolator<Real> interp(factory2, factory2);
   interp.Apply(field, interpolated);
 
-  auto interpolated_dof = interpolated.GetDOFHolderDataHost();
-  auto original_dof = field.GetDOFHolderDataHost();
+  auto interpolated_dof =
+    pcms::FlattenToRank1View(interpolated.GetDOFHolderDataHost());
+  auto original_dof = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   REQUIRE(interpolated_dof.size() == original_dof.size());
   for (int i = 0; i < static_cast<int>(interpolated_dof.size()); ++i) {
     REQUIRE_THAT(interpolated_dof[i],
@@ -137,8 +140,8 @@ TEST_CASE("Interpolator: construct once, apply twice with different data")
   interp.Apply(source, target);
 
   {
-    auto src_dof = source.GetDOFHolderDataHost();
-    auto tgt_dof = target.GetDOFHolderDataHost();
+    auto src_dof = pcms::FlattenToRank1View(source.GetDOFHolderDataHost());
+    auto tgt_dof = pcms::FlattenToRank1View(target.GetDOFHolderDataHost());
     REQUIRE(tgt_dof.size() == src_dof.size());
     for (int i = 0; i < static_cast<int>(tgt_dof.size()); ++i) {
       REQUIRE_THAT(tgt_dof[i], Catch::Matchers::WithinRel(src_dof[i], 0.001) ||
@@ -153,7 +156,7 @@ TEST_CASE("Interpolator: construct once, apply twice with different data")
   interp.Apply(source, target);
 
   {
-    auto tgt_dof = target.GetDOFHolderDataHost();
+    auto tgt_dof = pcms::FlattenToRank1View(target.GetDOFHolderDataHost());
     for (int i = 0; i < static_cast<int>(tgt_dof.size()); ++i) {
       REQUIRE_THAT(tgt_dof[i], Catch::Matchers::WithinAbs(7.0, 1E-10));
     }
diff --git a/test/test_mesh_intersection_field_transfer.cpp b/test/test_mesh_intersection_field_transfer.cpp
index 5a72b55f..bd34af54 100644
--- a/test/test_mesh_intersection_field_transfer.cpp
+++ b/test/test_mesh_intersection_field_transfer.cpp
@@ -208,15 +208,17 @@ TEST_CASE("OmegaHConservativeProjection matches conservative projection solver",
     });
 
   const auto intersections = pcms::intersectTargets(source_mesh, target_mesh);
-  const auto expected = pcms::solveGalerkinProjection(
-    target_mesh, source_mesh, intersections,
-    make_omega_h_reals(source.GetDOFHolderDataHost()));
+  const auto expected =
+    pcms::solveGalerkinProjection(target_mesh, source_mesh, intersections,
+                                  make_omega_h_reals(pcms::FlattenToRank1View(
+                                    source.GetDOFHolderDataHost())));
   const auto expected_h = Omega_h::HostRead<Omega_h::Real>(expected);
 
   pcms::OmegaHConservativeProjection projection(source_space, target_space);
   projection.Apply(source, target);
 
-  const auto target_values = target.GetDOFHolderDataHost();
+  const auto target_values =
+    pcms::FlattenToRank1View(target.GetDOFHolderDataHost());
   REQUIRE(static_cast<Omega_h::LO>(target_values.size()) ==
           target_mesh.nverts());
   for (Omega_h::LO i = 0; i < target_mesh.nverts(); ++i) {
diff --git a/test/test_omega_h_lagrange_field.cpp b/test/test_omega_h_lagrange_field.cpp
index 96da54f6..259a0165 100644
--- a/test/test_omega_h_lagrange_field.cpp
+++ b/test/test_omega_h_lagrange_field.cpp
@@ -40,7 +40,7 @@ TEST_CASE("OmegaHLagrangeLayout order-1 properties")
   REQUIRE(layout.IsDistributed()); // always true for Omega_h mesh layouts
 
   // DOF holder coordinates should match vertex coordinates
-  auto coords_device = layout.GetDOFHolderCoordinates().GetCoordinates();
+  auto coords_device = layout.GetDOFHolderCoordinates().GetValues();
   int nverts = mesh.nents(0);
   auto coords_view =
     pcms::test::CopyCoordinatesToHost(coords_device, nverts, mesh.dim());
@@ -74,7 +74,7 @@ TEST_CASE("OmegaHLagrangeLayout order-0 properties")
   REQUIRE(layout.GetNumGlobalDofHolder() == mesh.nglobal_ents(mesh.dim()));
 
   // DOF holder coordinates should match element centroids
-  auto coords_device = layout.GetDOFHolderCoordinates().GetCoordinates();
+  auto coords_device = layout.GetDOFHolderCoordinates().GetValues();
   int nelems = mesh.nelems();
   auto coords_view =
     pcms::test::CopyCoordinatesToHost(coords_device, nelems, mesh.dim());
@@ -136,10 +136,10 @@ TEST_CASE("OmegaHLagrangeField order-1: set/get DOF data round-trip")
   for (int i = 0; i < n; ++i)
     data[i] = static_cast<Real>(i);
 
-  pcms::Rank1View<const Real, pcms::HostMemorySpace> view(data.data(), n);
+  pcms::Rank2View<const Real, pcms::HostMemorySpace> view(data.data(), n, 1);
   field.GetData().SetDOFHolderDataHost(view);
 
-  auto got = field.GetData().GetDOFHolderDataHost();
+  auto got = pcms::FlattenToRank1View(field.GetData().GetDOFHolderDataHost());
   REQUIRE(static_cast<int>(got.size()) == n);
   for (int i = 0; i < n; ++i)
     REQUIRE(got[i] == Catch::Approx(data[i]));
@@ -210,6 +210,30 @@ TEST_CASE("OmegaHLagrangeField order-1: serialize / deserialize round-trip")
   pcms::test::CheckSerializeDeserialize(*factory.GetLayout(), field.GetData());
 }
 
+TEST_CASE(
+  "OmegaHLagrangeField order-1: multi-component serialize / deserialize "
+  "round-trip")
+{
+  auto lib = Omega_h::Library{};
+  auto mesh = MakeBox2D(lib.world());
+  const int nc = 3;
+  // The MeshFields backend is scalar-only; multi-component needs OmegaH.
+  auto factory = pcms::LagrangeFunctionSpace::FromMesh(
+    mesh, 1, nc, pcms::CoordinateSystem::Cartesian, "global",
+    pcms::LagrangeFunctionSpace::Backend::OmegaH);
+  auto field = factory.CreateField<Real>(pcms::FieldMetadata{});
+
+  const int n = factory.GetLayout()->GetNumOwnedDofHolder();
+  std::vector<Real> data(static_cast<size_t>(n) * nc);
+  for (int i = 0; i < n; ++i)
+    for (int c = 0; c < nc; ++c)
+      data[static_cast<size_t>(i) * nc + c] = i + 0.25 * c;
+  field.GetData().SetDOFHolderDataHost(
+    pcms::Rank2View<const Real, pcms::HostMemorySpace>(data.data(), n, nc));
+
+  pcms::test::CheckSerializeDeserialize(*factory.GetLayout(), field.GetData());
+}
+
 // ---- Order-0 field tests ----------------------------------------------------
 
 TEST_CASE("OmegaHLagrangeField order-0: set/get DOF data round-trip")
@@ -222,10 +246,10 @@ TEST_CASE("OmegaHLagrangeField order-0: set/get DOF data round-trip")
 
   int n = factory.GetLayout()->GetNumOwnedDofHolder();
   std::vector<Real> data(n, 3.14);
-  pcms::Rank1View<const Real, pcms::HostMemorySpace> view(data.data(), n);
+  pcms::Rank2View<const Real, pcms::HostMemorySpace> view(data.data(), n, 1);
   field.GetData().SetDOFHolderDataHost(view);
 
-  auto got = field.GetData().GetDOFHolderDataHost();
+  auto got = pcms::FlattenToRank1View(field.GetData().GetDOFHolderDataHost());
   REQUIRE(static_cast<int>(got.size()) == n);
   for (int i = 0; i < n; ++i)
     REQUIRE(got[i] == Catch::Approx(3.14));
@@ -242,7 +266,8 @@ TEST_CASE("OmegaHLagrangeField order-0: constant field evaluation")
   const Real kValue = 42.0;
   int nelems = mesh.nelems();
   std::vector<Real> data(nelems, kValue);
-  pcms::Rank1View<const Real, pcms::HostMemorySpace> view(data.data(), nelems);
+  pcms::Rank2View<const Real, pcms::HostMemorySpace> view(data.data(), nelems,
+                                                          1);
   field.GetData().SetDOFHolderDataHost(view);
 
   auto pts = pcms::test::StandardEvalCoords2D();
@@ -275,7 +300,8 @@ TEST_CASE("OmegaHLagrangeField order-0: out-of-bounds FILL mode")
 
   int nelems = mesh.nelems();
   std::vector<Real> data(nelems, 1.0);
-  pcms::Rank1View<const Real, pcms::HostMemorySpace> view(data.data(), nelems);
+  pcms::Rank2View<const Real, pcms::HostMemorySpace> view(data.data(), nelems,
+                                                          1);
   field.GetData().SetDOFHolderDataHost(view);
 
   Real fill_value = -1.0;
@@ -295,7 +321,8 @@ TEST_CASE("OmegaHLagrangeField order-0: serialize / deserialize round-trip")
   std::vector<Real> data(nelems);
   for (int i = 0; i < nelems; ++i)
     data[i] = static_cast<Real>(i);
-  pcms::Rank1View<const Real, pcms::HostMemorySpace> view(data.data(), nelems);
+  pcms::Rank2View<const Real, pcms::HostMemorySpace> view(data.data(), nelems,
+                                                          1);
   field.GetData().SetDOFHolderDataHost(view);
 
   pcms::test::CheckSerializeDeserialize(*factory.GetLayout(), field.GetData());
diff --git a/test/test_point_evaluator.cpp b/test/test_point_evaluator.cpp
index ae6b9245..fa0fc2f5 100644
--- a/test/test_point_evaluator.cpp
+++ b/test/test_point_evaluator.cpp
@@ -213,8 +213,8 @@ TEST_CASE("FieldLayout: metadata queries")
 
   auto coords = layout->GetDOFHolderCoordinates();
   REQUIRE(coords.GetCoordinateSystem() == CoordinateSystem::Cartesian);
-  REQUIRE(coords.GetCoordinates().extent(0) > 0);
-  REQUIRE(coords.GetCoordinates().extent(1) == 2);
+  REQUIRE(coords.GetValues().extent(0) > 0);
+  REQUIRE(coords.GetValues().extent(1) == 2);
 }
 
 // ============================================================================
@@ -234,8 +234,8 @@ TEST_CASE("FieldData: layout metadata queries")
 
   auto coords = factory.GetLayout()->GetDOFHolderCoordinates();
   REQUIRE(coords.GetCoordinateSystem() == CoordinateSystem::Cartesian);
-  REQUIRE(coords.GetCoordinates().extent(0) > 0);
-  REQUIRE(coords.GetCoordinates().extent(1) == 2);
+  REQUIRE(coords.GetValues().extent(0) > 0);
+  REQUIRE(coords.GetValues().extent(1) == 2);
 }
 
 // ============================================================================
@@ -263,9 +263,10 @@ TEST_CASE("SimpleFieldData: set and get DOF holder data round-trip")
     data_in[i] = static_cast<Real>(i) * 0.5;
 
   field_data.GetData().SetDOFHolderDataHost(
-    pcms::Rank1View<const Real, pcms::HostMemorySpace>(data_in.data(), n));
+    pcms::Rank2View<const Real, pcms::HostMemorySpace>(data_in.data(), n, 1));
 
-  auto data_out = field_data.GetData().GetDOFHolderDataHost();
+  auto data_out =
+    pcms::FlattenToRank1View(field_data.GetData().GetDOFHolderDataHost());
   REQUIRE(static_cast<int>(data_out.size()) == n);
   for (int i = 0; i < n; ++i) {
     REQUIRE(data_out[i] == Catch::Approx(data_in[i]));
@@ -290,8 +291,8 @@ TEST_CASE("FieldLayout: MeshFields metadata queries")
   auto layout = factory.GetLayout();
   auto coords = layout->GetDOFHolderCoordinates();
   REQUIRE(coords.GetCoordinateSystem() == CoordinateSystem::Cartesian);
-  REQUIRE(coords.GetCoordinates().extent(0) > 0);
-  REQUIRE(coords.GetCoordinates().extent(1) == 2);
+  REQUIRE(coords.GetValues().extent(0) > 0);
+  REQUIRE(coords.GetValues().extent(1) == 2);
 }
 
 TEST_CASE("PointEvaluator: MeshFields order-1 linear evaluation")
diff --git a/test/test_polynomial_reconstruction_function_space.cpp b/test/test_polynomial_reconstruction_function_space.cpp
index 9e0dc810..1f3b675a 100644
--- a/test/test_polynomial_reconstruction_function_space.cpp
+++ b/test/test_polynomial_reconstruction_function_space.cpp
@@ -44,7 +44,7 @@ TEST_CASE(
   REQUIRE(layout->GetNumGlobalDofHolder() == 4);
   REQUIRE_FALSE(layout->IsDistributed());
 
-  auto dof_coords = layout->GetDOFHolderCoordinates().GetCoordinates();
+  auto dof_coords = layout->GetDOFHolderCoordinates().GetValues();
   auto dof_coords_host = pcms::test::CopyCoordinatesToHost(dof_coords, 4, 2);
 
   REQUIRE(static_cast<int>(dof_coords_host.extent(0)) == 4);
@@ -81,10 +81,11 @@ TEST_CASE("PolynomialReconstructionFunctionSpace point-cloud field set/get DOF "
                  .CreateField<Real>(pcms::FieldMetadata{});
 
   std::vector<Real> data{1.0, 2.0, 3.0, 4.0};
-  Rank1View<const Real, HostMemorySpace> data_view(data.data(), data.size());
+  Rank2View<const Real, HostMemorySpace> data_view(
+    data.data(), static_cast<LO>(data.size()), 1);
   field.GetData().SetDOFHolderDataHost(data_view);
 
-  auto got = field.GetData().GetDOFHolderDataHost();
+  auto got = pcms::FlattenToRank1View(field.GetData().GetDOFHolderDataHost());
   REQUIRE(got.size() == data.size());
   for (LO i = 0; i < static_cast<LO>(data.size()); ++i) {
     REQUIRE(got[i] == Catch::Approx(data[i]));
@@ -102,7 +103,8 @@ TEST_CASE("PolynomialReconstructionFunctionSpace point-cloud field serialize / "
   auto field = factory.CreateField<Real>(pcms::FieldMetadata{});
 
   std::vector<Real> data{5.0, 6.0, 7.0, 8.0};
-  Rank1View<const Real, HostMemorySpace> data_view(data.data(), data.size());
+  Rank2View<const Real, HostMemorySpace> data_view(
+    data.data(), static_cast<LO>(data.size()), 1);
   field.GetData().SetDOFHolderDataHost(data_view);
 
   pcms::test::CheckSerializeDeserialize(*factory.GetLayout(), field.GetData());
@@ -126,10 +128,11 @@ TEST_CASE("PolynomialReconstructionFunctionSpace field keeps layout alive "
   REQUIRE(point_cloud_layout->GetNumOwnedDofHolder() == 4);
 
   std::vector<Real> data{9.0, 10.0, 11.0, 12.0};
-  Rank1View<const Real, HostMemorySpace> data_view(data.data(), data.size());
+  Rank2View<const Real, HostMemorySpace> data_view(
+    data.data(), static_cast<LO>(data.size()), 1);
   field.SetDOFHolderDataHost(data_view);
 
-  auto got = field.GetDOFHolderDataHost();
+  auto got = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   REQUIRE(got[0] == Catch::Approx(9.0));
   REQUIRE(got[3] == Catch::Approx(12.0));
 }
diff --git a/test/test_polynomial_reconstruction_mls_evaluation.cpp b/test/test_polynomial_reconstruction_mls_evaluation.cpp
index 3aef08d8..8b979e68 100644
--- a/test/test_polynomial_reconstruction_mls_evaluation.cpp
+++ b/test/test_polynomial_reconstruction_mls_evaluation.cpp
@@ -334,8 +334,8 @@ TEST_CASE("PolynomialReconstructionFunctionSpace MLS: radius option is "
     coords_view, CoordinateSystem::Cartesian, opts);
   auto field = fs.CreateField<Real>(pcms::FieldMetadata{});
   std::vector<Real> dof_values{1.0, 5.0};
-  field.GetData().SetDOFHolderDataHost(Rank1View<const Real, HostMemorySpace>(
-    dof_values.data(), dof_values.size()));
+  field.GetData().SetDOFHolderDataHost(Rank2View<const Real, HostMemorySpace>(
+    dof_values.data(), static_cast<pcms::LO>(dof_values.size()), 1));
 
   std::vector<Real> pts{0.0, 0.0};
   auto device_coords =
@@ -364,7 +364,7 @@ TEST_CASE("PolynomialReconstructionFunctionSpace MLS: 3D point clouds preserve "
   auto fs = pcms::PolynomialReconstructionFunctionSpace::Create(
     coords_view, CoordinateSystem::Cartesian, DefaultTestOptions3D());
   auto layout_coords_device =
-    fs.GetLayout()->GetDOFHolderCoordinates().GetCoordinates();
+    fs.GetLayout()->GetDOFHolderCoordinates().GetValues();
   auto layout_coords =
     pcms::test::CopyCoordinatesToHost(layout_coords_device, 27, 3);
   REQUIRE(static_cast<int>(layout_coords.extent(1)) == 3);
@@ -382,8 +382,8 @@ TEST_CASE("PolynomialReconstructionFunctionSpace MLS: 3D point clouds preserve "
     const Real z = src[3 * i + 2];
     dof_values[i] = x + 2.0 * y + 3.0 * z;
   }
-  field.GetData().SetDOFHolderDataHost(Rank1View<const Real, HostMemorySpace>(
-    dof_values.data(), dof_values.size()));
+  field.GetData().SetDOFHolderDataHost(Rank2View<const Real, HostMemorySpace>(
+    dof_values.data(), static_cast<pcms::LO>(dof_values.size()), 1));
 
   std::vector<Real> pts{0.5, 0.5, 0.25, 0.5, 0.5, 0.75};
   auto device_coords =
diff --git a/test/test_proxy_coupling.cpp b/test/test_proxy_coupling.cpp
index 759cd8d3..455fc183 100644
--- a/test/test_proxy_coupling.cpp
+++ b/test/test_proxy_coupling.cpp
@@ -35,7 +35,8 @@ void initializeFieldWithGids(const pcms::FieldLayout& layout,
     Kokkos::RangePolicy<pcms::HostMemorySpace::execution_space>(0, n),
     [=](int i) { ids[i] = gids[i] * multiplier; });
 
-  field->SetDOFHolderDataHost(pcms::make_const_array_view(ids));
+  field->SetDOFHolderDataHost(
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(ids.data(), n, 1));
 }
 
 bool validateField(const pcms::FieldLayout& layout,
@@ -44,7 +45,7 @@ bool validateField(const pcms::FieldLayout& layout,
                    pcms::Real multiplier = 1.0)
 {
   auto gids = layout.GetGidsHost();
-  auto copied_array = field->GetDOFHolderDataHost();
+  auto copied_array = pcms::FlattenToRank1View(field->GetDOFHolderDataHost());
   auto owned = layout.GetOwnedHost();
   const auto n = layout.GetNumOwnedDofHolder();
 
diff --git a/test/test_uniform_grid_field.cpp b/test/test_uniform_grid_field.cpp
index acbbe1a8..b179167d 100644
--- a/test/test_uniform_grid_field.cpp
+++ b/test/test_uniform_grid_field.cpp
@@ -70,8 +70,8 @@ void VerifyUniformGridFieldValues(
   for (int j = 0; j <= grid.divisions[1]; ++j) {
     for (int i = 0; i <= grid.divisions[0]; ++i) {
       int vertex_id = j * (grid.divisions[0] + 1) + i;
-      pcms::Real x = ug_coords.GetCoordinates()(vertex_id, 0);
-      pcms::Real y = ug_coords.GetCoordinates()(vertex_id, 1);
+      pcms::Real x = ug_coords.GetValues()(vertex_id, 0);
+      pcms::Real y = ug_coords.GetValues()(vertex_id, 1);
       pcms::Real expected = x + 2.0 * y;
       pcms::Real actual = ug_field_data[vertex_id];
       REQUIRE(std::abs(expected - actual) <= 1e-10);
@@ -83,7 +83,7 @@ void VerifyUniformGridFieldValues(
 void VerifyMaskFieldValues(const pcms::UniformGrid<2>& grid,
                            const pcms::Field<pcms::Real>& mask_field)
 {
-  auto mask_data = mask_field.GetDOFHolderDataHost();
+  auto mask_data = pcms::FlattenToRank1View(mask_field.GetDOFHolderDataHost());
   for (int j = 0; j <= grid.divisions[1]; ++j) {
     for (int i = 0; i <= grid.divisions[0]; ++i) {
       int vertex_id = j * (grid.divisions[0] + 1) + i;
@@ -131,7 +131,7 @@ TEST_CASE("UniformGrid order-0 field creation and evaluation")
   REQUIRE(layout->GetOrder() == 0);
   REQUIRE(layout->GetNumOwnedDofHolder() == 4);
 
-  auto coords_device = layout->GetDOFHolderCoordinates().GetCoordinates();
+  auto coords_device = layout->GetDOFHolderCoordinates().GetValues();
   auto coords = pcms::test::CopyCoordinatesToHost(coords_device, 4, 2);
 
   REQUIRE(coords(0, 0) == Catch::Approx(2.5));
@@ -141,8 +141,8 @@ TEST_CASE("UniformGrid order-0 field creation and evaluation")
 
   std::vector<pcms::Real> data = {1.0, 2.0, 3.0, 4.0};
   field.SetDOFHolderDataHost(
-    pcms::Rank1View<const pcms::Real, pcms::HostMemorySpace>(data.data(),
-                                                             data.size()));
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(
+      data.data(), static_cast<pcms::LO>(data.size()), 1));
 
   std::vector<pcms::Real> eval_coords = {1.0, 1.0, 9.0, 1.0,
                                          1.0, 9.0, 9.0, 9.0};
@@ -186,10 +186,10 @@ TEST_CASE("UniformGrid field data operations", "[uniform_grid_field]")
     data[i] = static_cast<pcms::Real>(i);
 
   field.SetDOFHolderDataHost(
-    pcms::Rank1View<const pcms::Real, pcms::HostMemorySpace>(data.data(),
-                                                             data.size()));
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(
+      data.data(), static_cast<pcms::LO>(data.size()), 1));
 
-  auto retrieved = field.GetDOFHolderDataHost();
+  auto retrieved = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   REQUIRE(retrieved.size() == 25);
   for (size_t i = 0; i < 25; ++i)
     REQUIRE(retrieved[i] == static_cast<pcms::Real>(i));
@@ -222,8 +222,8 @@ TEST_CASE("UniformGrid field evaluation - piecewise constant")
     3.0, 3.5, 4.0  // v6, v7, v8 (top row, y=10)
   };
   field.SetDOFHolderDataHost(
-    pcms::Rank1View<const pcms::Real, pcms::HostMemorySpace>(data.data(),
-                                                             data.size()));
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(
+      data.data(), static_cast<pcms::LO>(data.size()), 1));
 
   std::vector<pcms::Real> eval_coords = {
     2.5, 2.5, // Cell 0 center
@@ -276,8 +276,8 @@ TEST_CASE("UniformGrid field serialization")
     data[i] = static_cast<pcms::Real>(i * 10);
 
   field.SetDOFHolderDataHost(
-    pcms::Rank1View<const pcms::Real, pcms::HostMemorySpace>(data.data(),
-                                                             data.size()));
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(
+      data.data(), static_cast<pcms::LO>(data.size()), 1));
 
   pcms::test::CheckSerializeDeserialize(field);
 }
@@ -305,14 +305,14 @@ TEST_CASE("UniformGrid field copy")
     grid, 1, pcms::CoordinateSystem::Cartesian);
   auto field = factory.CreateField<pcms::Real>(pcms::FieldMetadata{});
   field.SetDOFHolderDataHost(
-    pcms::Rank1View<const pcms::Real, pcms::HostMemorySpace>(data.data(),
-                                                             data.size()));
+    pcms::Rank2View<const pcms::Real, pcms::HostMemorySpace>(
+      data.data(), static_cast<pcms::LO>(data.size()), 1));
 
   auto field2 = factory.CreateField<pcms::Real>(pcms::FieldMetadata{});
   pcms::Copy<pcms::Real> copy(factory, factory);
   copy.Apply(field, field2);
 
-  auto copied_data = field2.GetDOFHolderDataHost();
+  auto copied_data = pcms::FlattenToRank1View(field2.GetDOFHolderDataHost());
   REQUIRE(copied_data.size() == data.size());
   for (size_t i = 0; i < data.size(); ++i)
     REQUIRE(copied_data[i] == data[i]);
@@ -343,13 +343,14 @@ TEST_CASE("Transfer from OmegaH field to UniformGrid field")
   pcms::Interpolator<pcms::Real> interp(omega_h_factory, ug_factory);
   interp.Apply(omega_h_field, ug_field);
 
-  auto transferred_data = ug_field.GetDOFHolderDataHost();
+  auto transferred_data =
+    pcms::FlattenToRank1View(ug_field.GetDOFHolderDataHost());
   auto ug_coords = ug_factory.GetLayout()->GetDOFHolderCoordinates();
   int num_ug_nodes = ug_factory.GetLayout()->GetNumOwnedDofHolder();
 
   // set up_coords to host
-  auto ug_coords_host = pcms::test::CopyCoordinatesToHost(
-    ug_coords.GetCoordinates(), num_ug_nodes, 2);
+  auto ug_coords_host =
+    pcms::test::CopyCoordinatesToHost(ug_coords.GetValues(), num_ug_nodes, 2);
 
   for (int i = 0; i < num_ug_nodes; ++i) {
     pcms::Real x = ug_coords_host(i, 0);
@@ -371,7 +372,7 @@ TEST_CASE("Create binary field from uniform grid")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{5, 5});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 36); // (5+1) * (5+1) = 36 vertices
 
@@ -390,7 +391,7 @@ TEST_CASE("Create binary field from uniform grid")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{10, 8});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 99); // (10+1) * (8+1) = 99 vertices
 
@@ -408,7 +409,7 @@ TEST_CASE("Create binary field from uniform grid")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{10, 10});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     for (size_t i = 0; i < field_data.size(); ++i) {
       REQUIRE((field_data[i] == 0.0 || field_data[i] == 1.0));
@@ -426,7 +427,7 @@ TEST_CASE("Create binary field from uniform grid")
     grid.divisions = {10, 10};
 
     auto [layout, field] = pcms::CreateUniformGridBinaryField<2>(mesh, grid);
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 121); // (10+1) * (10+1) = 121 vertices
 
@@ -460,7 +461,7 @@ TEST_CASE("Create binary field from uniform grid")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{20, 20});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 441); // (20+1) * (20+1) = 441 vertices
 
@@ -481,7 +482,7 @@ TEST_CASE("Create binary field from uniform grid")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{30, 10});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 341); // (30+1) * (10+1) = 341 vertices
 
@@ -508,7 +509,7 @@ TEST_CASE("Binary field integration with grid methods")
     auto grid = CreateUniformGridFromMesh<2>(mesh, std::array{8, 8});
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{8, 8});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     // Get field value for a specific vertex (middle vertex at i=4, j=4)
     pcms::LO vertex_id = 4 * 9 + 4;
@@ -528,7 +529,7 @@ TEST_CASE("Binary field integration with grid methods")
     auto grid = CreateUniformGridFromMesh<2>(mesh, std::array{10, 10});
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{10, 10});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     int q1 = 0, q2 = 0, q3 = 0, q4 = 0;
 
@@ -583,7 +584,7 @@ TEST_CASE("Performance and edge cases")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{50, 50});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 2601); // (50+1) * (50+1) = 2601 vertices
 
@@ -600,7 +601,7 @@ TEST_CASE("Performance and edge cases")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{2, 2});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 9); // (2+1) * (2+1) = 9 vertices
 
@@ -617,7 +618,7 @@ TEST_CASE("Performance and edge cases")
 
     auto [layout, field] =
       pcms::CreateUniformGridBinaryField<2>(mesh, std::array{25, 10});
-    auto field_data = field.GetDOFHolderDataHost();
+    auto field_data = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
 
     REQUIRE(field_data.size() == 286); // (25+1) * (10+1) = 286 vertices
 
@@ -655,7 +656,7 @@ TEST_CASE("UniformGrid workflow")
   pcms::Interpolator<pcms::Real> interp(omega_h_factory, ug_factory);
   interp.Apply(omega_h_field, ug_field);
   auto ug_coords_device_view =
-    ug_factory.GetLayout()->GetDOFHolderCoordinates().GetCoordinates();
+    ug_factory.GetLayout()->GetDOFHolderCoordinates().GetValues();
   auto ug_coords_host_view =
     pcms::test::CopyCoordinatesToHost(ug_coords_device_view, 25, 2);
 
@@ -664,7 +665,7 @@ TEST_CASE("UniformGrid workflow")
     "", 25);
   Kokkos::parallel_for(
     "CopyFieldDataToView", 25, KOKKOS_LAMBDA(int i) {
-      ug_field_data_device_view(i) = ug_field_data_device(i);
+      ug_field_data_device_view(i) = ug_field_data_device(i, 0);
     });
   auto ug_field_data_host_view =
     Kokkos::View<pcms::Real*, pcms::HostMemorySpace>("", 25);
diff --git a/test/test_xgc_field_data.cpp b/test/test_xgc_field_data.cpp
index de3f8411..234bd2d1 100644
--- a/test/test_xgc_field_data.cpp
+++ b/test/test_xgc_field_data.cpp
@@ -94,7 +94,7 @@ TEST_CASE("XGC FieldData serializer preserves inactive entries")
   serializer.Deserialize(field, *layout, pcms::make_const_array_view(buffer),
                          pcms::make_const_array_view(permutation));
 
-  auto after = field.GetDOFHolderDataHost();
+  auto after = pcms::FlattenToRank1View(field.GetDOFHolderDataHost());
   for (int i = 0; i < data_size; ++i) {
     if (owned[i]) {
       REQUIRE(after[i] == Catch::Approx(original[i] + 100.0));