rtx-cfd legacy GPU: per-block buffer mutability; the D3Q19 getters extract before borrowing

crates/specialized/rtx-cfd/src/solvers/incompressible/piso_gpu.rs

@@ -284,7 +284,7 @@ impl PisoGpuSolver {
         let iterations = {
             let buffers = self
                 .gpu_buffers
-                .as_ref()
+                .as_mut()
                 .ok_or_else(|| CfdError::gpu_error("GPU buffers not initialized"))?;
             self.poisson_kernel.solve_2d(
                 &mut buffers.pressure_correction,
@@ -303,7 +303,7 @@ impl PisoGpuSolver {
         {
             let buffers = self
                 .gpu_buffers
-                .as_ref()
+                .as_mut()
                 .ok_or_else(|| CfdError::gpu_error("GPU buffers not initialized"))?;
             self.matrix_ops_kernel
                 .axpy(1.0, &buffers.pressure_correction, &mut buffers.p)?;

crates/specialized/rtx-cfd/src/solvers/lbm/d3q19_gpu.rs

@@ -731,12 +731,12 @@ __global__ void d3q19_bounce_back_boundaries(
 
     /// Calculate total mass on GPU
     pub fn gpu_total_mass(&mut self) -> CfdResult<f64> {
+        self.gpu_extract_macroscopic_variables()?;
         let buffers = self
             .gpu_buffers
             .as_ref()
             .ok_or_else(|| CfdError::gpu_error("GPU buffers not initialized"))?;
 
-        self.gpu_extract_macroscopic_variables()?;
         let density_host = self.kernel_manager.copy_from_device(&buffers.density)?;
         let total_mass: f32 = density_host.iter().sum();
         Ok(total_mass as f64)
@@ -744,12 +744,12 @@ __global__ void d3q19_bounce_back_boundaries(
 
     /// Calculate kinetic energy on GPU
     pub fn gpu_kinetic_energy(&mut self) -> CfdResult<f64> {
+        self.gpu_extract_macroscopic_variables()?;
         let buffers = self
             .gpu_buffers
             .as_ref()
             .ok_or_else(|| CfdError::gpu_error("GPU buffers not initialized"))?;
 
-        self.gpu_extract_macroscopic_variables()?;
         let density_host = self.kernel_manager.copy_from_device(&buffers.density)?;
         let velocity_x_host = self.kernel_manager.copy_from_device(&buffers.velocity_x)?;
         let velocity_y_host = self.kernel_manager.copy_from_device(&buffers.velocity_y)?;