diff --git a/crates/specialized/rtx-cfd/src/solvers/incompressible/piso_gpu.rs b/crates/specialized/rtx-cfd/src/solvers/incompressible/piso_gpu.rs
index e59e40c..2f078ed 100644
--- a/crates/specialized/rtx-cfd/src/solvers/incompressible/piso_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/solvers/incompressible/piso_gpu.rs
@@ -377,21 +377,23 @@ impl PisoGpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.mass_source)
-                .arg(u)
-                .arg(v)
-                .arg(&(density / dt))
-                .arg(&dx)
-                .arg(&dy)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Divergence kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.mass_source)
+                    .arg(u)
+                    .arg(v)
+                    .arg(&(density / dt))
+                    .arg(&dx)
+                    .arg(&dy)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Divergence kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -460,21 +462,23 @@ impl PisoGpuSolver {
         let correction_factor = -dt / density;
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.pressure_correction)
-                .arg(&mut buffers.u_correction.clone())
-                .arg(&mut buffers.v_correction.clone())
-                .arg(&correction_factor)
-                .arg(&dx)
-                .arg(&dy)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Gradient kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.pressure_correction)
+                    .arg(&mut buffers.u_correction.clone())
+                    .arg(&mut buffers.v_correction.clone())
+                    .arg(&correction_factor)
+                    .arg(&dx)
+                    .arg(&dy)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Gradient kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -644,6 +648,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = PisoParameters::default();
@@ -702,6 +707,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = PisoParameters::default();
diff --git a/crates/specialized/rtx-cfd/src/solvers/incompressible/simple_gpu.rs b/crates/specialized/rtx-cfd/src/solvers/incompressible/simple_gpu.rs
index 3bdd7ca..b09cb0b 100644
--- a/crates/specialized/rtx-cfd/src/solvers/incompressible/simple_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/solvers/incompressible/simple_gpu.rs
@@ -499,6 +499,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = SimpleParameters::new();
diff --git a/crates/specialized/rtx-cfd/src/solvers/lbm/d2q9_gpu.rs b/crates/specialized/rtx-cfd/src/solvers/lbm/d2q9_gpu.rs
index 1052da7..be8f798 100644
--- a/crates/specialized/rtx-cfd/src/solvers/lbm/d2q9_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/solvers/lbm/d2q9_gpu.rs
@@ -215,20 +215,22 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut f.clone())
-                .arg(density)
-                .arg(velocity_x)
-                .arg(velocity_y)
-                .arg(&(self.cs2 as f32))
-                .arg(&(nx as i32))
-                .arg(&(ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Equilibrium init kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut f.clone())
+                    .arg(density)
+                    .arg(velocity_x)
+                    .arg(velocity_y)
+                    .arg(&(self.cs2 as f32))
+                    .arg(&(nx as i32))
+                    .arg(&(ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Equilibrium init kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -281,18 +283,20 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.f.clone())
-                .arg(&buffers.f_eq)
-                .arg(&(self.omega as f32))
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Collision kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.f.clone())
+                    .arg(&buffers.f_eq)
+                    .arg(&(self.omega as f32))
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Collision kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -324,17 +328,19 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.f_temp.clone())
-                .arg(&buffers.f)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Streaming kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.f_temp.clone())
+                    .arg(&buffers.f)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Streaming kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         // Swap buffers: f = f_temp
@@ -368,22 +374,24 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.density.clone())
-                .arg(&mut buffers.velocity_x.clone())
-                .arg(&mut buffers.velocity_y.clone())
-                .arg(&buffers.f)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!(
-                        "Macroscopic variables kernel launch failed: {}",
-                        e
-                    ))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.density.clone())
+                    .arg(&mut buffers.velocity_x.clone())
+                    .arg(&mut buffers.velocity_y.clone())
+                    .arg(&buffers.f)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!(
+                            "Macroscopic variables kernel launch failed: {}",
+                            e
+                        ))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -415,19 +423,21 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.f_eq.clone())
-                .arg(&buffers.density)
-                .arg(&buffers.velocity_x)
-                .arg(&buffers.velocity_y)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Equilibrium kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.f_eq.clone())
+                    .arg(&buffers.density)
+                    .arg(&buffers.velocity_x)
+                    .arg(&buffers.velocity_y)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Equilibrium kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -459,16 +469,18 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.f.clone())
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Boundary kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.f.clone())
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Boundary kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -513,22 +525,24 @@ impl D2Q9GpuSolver {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.f.clone())
-                .arg(&mut buffers.density.clone())
-                .arg(&mut buffers.velocity_x.clone())
-                .arg(&mut buffers.velocity_y.clone())
-                .arg(&(density as f32))
-                .arg(&(velocity.x as f32))
-                .arg(&(velocity.y as f32))
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Initialization kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.f.clone())
+                    .arg(&mut buffers.density.clone())
+                    .arg(&mut buffers.velocity_x.clone())
+                    .arg(&mut buffers.velocity_y.clone())
+                    .arg(&(density as f32))
+                    .arg(&(velocity.x as f32))
+                    .arg(&(velocity.y as f32))
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Initialization kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
diff --git a/crates/specialized/rtx-cfd/src/solvers/lbm/d3q19_gpu.rs b/crates/specialized/rtx-cfd/src/solvers/lbm/d3q19_gpu.rs
index a55a5d6..880f800 100644
--- a/crates/specialized/rtx-cfd/src/solvers/lbm/d3q19_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/solvers/lbm/d3q19_gpu.rs
@@ -395,17 +395,19 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.f.clone())
-            .arg(&buffers.f_eq)
-            .arg(&(omega as f32))
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| CfdError::gpu_error(&format!("Collision kernel launch failed: {}", e)))?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.f.clone())
+                .arg(&buffers.f_eq)
+                .arg(&(omega as f32))
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| CfdError::gpu_error(&format!("Collision kernel launch failed: {}", e)))?;
+        }
 
         self.kernel_manager.synchronize()?;
         Ok(())
@@ -436,16 +438,18 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.f_temp.clone())
-            .arg(&buffers.f)
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| CfdError::gpu_error(&format!("Streaming kernel launch failed: {}", e)))?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.f_temp.clone())
+                .arg(&buffers.f)
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| CfdError::gpu_error(&format!("Streaming kernel launch failed: {}", e)))?;
+        }
 
         // Swap buffers: f = f_temp
         // In real implementation, would swap buffer pointers
@@ -483,24 +487,26 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.density.clone())
-            .arg(&mut buffers.velocity_x.clone())
-            .arg(&mut buffers.velocity_y.clone())
-            .arg(&mut buffers.velocity_z.clone())
-            .arg(&buffers.f)
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| {
-                CfdError::gpu_error(&format!(
-                    "Macroscopic variables kernel launch failed: {}",
-                    e
-                ))
-            })?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.density.clone())
+                .arg(&mut buffers.velocity_x.clone())
+                .arg(&mut buffers.velocity_y.clone())
+                .arg(&mut buffers.velocity_z.clone())
+                .arg(&buffers.f)
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| {
+                    CfdError::gpu_error(&format!(
+                        "Macroscopic variables kernel launch failed: {}",
+                        e
+                    ))
+                })?;
+        }
 
         self.kernel_manager.synchronize()?;
         Ok(())
@@ -531,21 +537,23 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.f_eq.clone())
-            .arg(&buffers.density)
-            .arg(&buffers.velocity_x)
-            .arg(&buffers.velocity_y)
-            .arg(&buffers.velocity_z)
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| {
-                CfdError::gpu_error(&format!("Equilibrium kernel launch failed: {}", e))
-            })?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.f_eq.clone())
+                .arg(&buffers.density)
+                .arg(&buffers.velocity_x)
+                .arg(&buffers.velocity_y)
+                .arg(&buffers.velocity_z)
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| {
+                    CfdError::gpu_error(&format!("Equilibrium kernel launch failed: {}", e))
+                })?;
+        }
 
         self.kernel_manager.synchronize()?;
         Ok(())
@@ -575,15 +583,17 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.f.clone())
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| CfdError::gpu_error(&format!("Boundary kernel launch failed: {}", e)))?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.f.clone())
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| CfdError::gpu_error(&format!("Boundary kernel launch failed: {}", e)))?;
+        }
 
         self.kernel_manager.synchronize()?;
         Ok(())
@@ -628,21 +638,23 @@ __global__ void d3q19_bounce_back_boundaries(
             shared_mem_bytes: 0,
         };
 
-        self.kernel_manager
-            .stream()
-            .launch_builder(&func)
-            .arg(&mut buffers.f.clone())
-            .arg(&(density as f32))
-            .arg(&(velocity.x as f32))
-            .arg(&(velocity.y as f32))
-            .arg(&(velocity.z as f32))
-            .arg(&(buffers.nx as i32))
-            .arg(&(buffers.ny as i32))
-            .arg(&(buffers.nz as i32))
-            .launch(config)
-            .map_err(|e| {
-                CfdError::gpu_error(&format!("Initialization kernel launch failed: {}", e))
-            })?;
+        unsafe {
+            self.kernel_manager
+                .stream()
+                .launch_builder(&func)
+                .arg(&mut buffers.f.clone())
+                .arg(&(density as f32))
+                .arg(&(velocity.x as f32))
+                .arg(&(velocity.y as f32))
+                .arg(&(velocity.z as f32))
+                .arg(&(buffers.nx as i32))
+                .arg(&(buffers.ny as i32))
+                .arg(&(buffers.nz as i32))
+                .launch(config)
+                .map_err(|e| {
+                    CfdError::gpu_error(&format!("Initialization kernel launch failed: {}", e))
+                })?;
+        }
 
         self.kernel_manager.synchronize()?;
         Ok(())
@@ -771,6 +783,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = D3Q19Parameters::default();
@@ -799,6 +812,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = D3Q19Parameters::default();
@@ -839,6 +853,7 @@ mod tests {
             viscosity: 0.01,
             density: 1.0,
             device_id: 0,
+            ..CfdConfig::default()
         };
 
         let params = D3Q19Parameters::default();
diff --git a/crates/specialized/rtx-cfd/src/turbulence/k_epsilon_gpu.rs b/crates/specialized/rtx-cfd/src/turbulence/k_epsilon_gpu.rs
index 7bfee41..6a50f82 100644
--- a/crates/specialized/rtx-cfd/src/turbulence/k_epsilon_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/turbulence/k_epsilon_gpu.rs
@@ -68,7 +68,7 @@ impl KEpsilonGpuModel {
 
         Ok(Self {
             kernel_manager,
-            constants: KEpsilonConstants::default(),
+            constants: KEpsilonConstants::standard(),
             k_epsilon_module: None,
             gpu_buffers: None,
         })
@@ -89,7 +89,7 @@ impl KEpsilonGpuModel {
     fn load_k_epsilon_kernels(&mut self) -> CfdResult<()> {
         // In real implementation, would compile or load PTX for k-epsilon kernels
         // For now, load from kernel manager
-        self.k_epsilon_module = Some(self.kernel_manager.get_module("k_epsilon_kernels")?);
+        self.k_epsilon_module = Some(self.kernel_manager.get_module("k_epsilon_kernels")?.clone());
         Ok(())
     }
 
@@ -140,12 +140,12 @@ impl KEpsilonGpuModel {
         let epsilon_host = vec![epsilon_init; n];
 
         self.kernel_manager
-            .stream
-            .memcpy_htod(&mut buffers.k, &k_host)
+            .stream()
+            .memcpy_htod(&k_host, &mut buffers.k)
             .map_err(|e| CfdError::gpu_error(&format!("Failed to copy k to GPU: {}", e)))?;
         self.kernel_manager
-            .stream
-            .memcpy_htod(&mut buffers.epsilon, &epsilon_host)
+            .stream()
+            .memcpy_htod(&epsilon_host, &mut buffers.epsilon)
             .map_err(|e| CfdError::gpu_error(&format!("Failed to copy epsilon to GPU: {}", e)))?;
 
         self.kernel_manager.synchronize()?;
@@ -188,31 +188,33 @@ impl KEpsilonGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(u)
-                .arg(v)
-                .arg(w)
-                .arg(&mut buffers.dudx)
-                .arg(&mut buffers.dudy)
-                .arg(&mut buffers.dudz)
-                .arg(&mut buffers.dvdx)
-                .arg(&mut buffers.dvdy)
-                .arg(&mut buffers.dvdz)
-                .arg(&mut buffers.dwdx)
-                .arg(&mut buffers.dwdy)
-                .arg(&mut buffers.dwdz)
-                .arg(&dx)
-                .arg(&dy)
-                .arg(&dz)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Velocity gradients kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(u)
+                    .arg(v)
+                    .arg(w)
+                    .arg(&mut buffers.dudx)
+                    .arg(&mut buffers.dudy)
+                    .arg(&mut buffers.dudz)
+                    .arg(&mut buffers.dvdx)
+                    .arg(&mut buffers.dvdy)
+                    .arg(&mut buffers.dvdz)
+                    .arg(&mut buffers.dwdx)
+                    .arg(&mut buffers.dwdy)
+                    .arg(&mut buffers.dwdz)
+                    .arg(&dx)
+                    .arg(&dy)
+                    .arg(&dz)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Velocity gradients kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -247,20 +249,22 @@ impl KEpsilonGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.k)
-                .arg(&buffers.epsilon)
-                .arg(&mut buffers.eddy_viscosity)
-                .arg(&self.constants.c_mu)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Eddy viscosity kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.k)
+                    .arg(&buffers.epsilon)
+                    .arg(&mut buffers.eddy_viscosity)
+                    .arg(&self.constants.c_mu)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Eddy viscosity kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -293,27 +297,29 @@ impl KEpsilonGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.eddy_viscosity)
-                .arg(&buffers.dudx)
-                .arg(&buffers.dudy)
-                .arg(&buffers.dudz)
-                .arg(&buffers.dvdx)
-                .arg(&buffers.dvdy)
-                .arg(&buffers.dvdz)
-                .arg(&buffers.dwdx)
-                .arg(&buffers.dwdy)
-                .arg(&buffers.dwdz)
-                .arg(&mut buffers.production)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Production kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.eddy_viscosity)
+                    .arg(&buffers.dudx)
+                    .arg(&buffers.dudy)
+                    .arg(&buffers.dudz)
+                    .arg(&buffers.dvdx)
+                    .arg(&buffers.dvdy)
+                    .arg(&buffers.dvdz)
+                    .arg(&buffers.dwdx)
+                    .arg(&buffers.dwdy)
+                    .arg(&buffers.dwdz)
+                    .arg(&mut buffers.production)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Production kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -348,22 +354,24 @@ impl KEpsilonGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.k)
-                .arg(&buffers.epsilon)
-                .arg(&buffers.production)
-                .arg(&mut buffers.source_epsilon)
-                .arg(&self.constants.c_e1)
-                .arg(&self.constants.c_e2)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Epsilon source kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.k)
+                    .arg(&buffers.epsilon)
+                    .arg(&buffers.production)
+                    .arg(&mut buffers.source_epsilon)
+                    .arg(&self.constants.c1_epsilon)
+                    .arg(&self.constants.c2_epsilon)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Epsilon source kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -400,25 +408,27 @@ impl KEpsilonGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.k)
-                .arg(&mut buffers.epsilon)
-                .arg(&buffers.k_old)
-                .arg(&buffers.epsilon_old)
-                .arg(&buffers.production)
-                .arg(&buffers.source_epsilon)
-                .arg(&buffers.eddy_viscosity)
-                .arg(&dt)
-                .arg(&nu)
-                .arg(&self.constants.sigma_k)
-                .arg(&self.constants.sigma_e)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| CfdError::gpu_error(&format!("Update kernel launch failed: {}", e)))?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.k)
+                    .arg(&mut buffers.epsilon)
+                    .arg(&buffers.k_old)
+                    .arg(&buffers.epsilon_old)
+                    .arg(&buffers.production)
+                    .arg(&buffers.source_epsilon)
+                    .arg(&buffers.eddy_viscosity)
+                    .arg(&dt)
+                    .arg(&nu)
+                    .arg(&self.constants.sigma_k)
+                    .arg(&self.constants.sigma_epsilon)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| CfdError::gpu_error(&format!("Update kernel launch failed: {}", e)))?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -436,7 +446,7 @@ impl KEpsilonGpuModel {
         let mut eddy_viscosity_host = vec![0.0f32; n];
 
         self.kernel_manager
-            .stream
+            .stream()
             .memcpy_dtoh(&mut eddy_viscosity_host, &buffers.eddy_viscosity)
             .map_err(|e| {
                 CfdError::gpu_error(&format!("Failed to copy eddy viscosity from GPU: {}", e))
diff --git a/crates/specialized/rtx-cfd/src/turbulence/smagorinsky_gpu.rs b/crates/specialized/rtx-cfd/src/turbulence/smagorinsky_gpu.rs
index 23b287b..c47dfb6 100644
--- a/crates/specialized/rtx-cfd/src/turbulence/smagorinsky_gpu.rs
+++ b/crates/specialized/rtx-cfd/src/turbulence/smagorinsky_gpu.rs
@@ -60,7 +60,7 @@ impl SmagorinskyGpuModel {
 
         Ok(Self {
             kernel_manager,
-            constants: SmagorinskyConstants::default(),
+            constants: SmagorinskyConstants::standard(),
             smagorinsky_module: None,
             gpu_buffers: None,
         })
@@ -80,7 +80,7 @@ impl SmagorinskyGpuModel {
     /// Load and compile Smagorinsky CUDA kernels
     fn load_smagorinsky_kernels(&mut self) -> CfdResult<()> {
         // In real implementation, would compile or load PTX for Smagorinsky kernels
-        self.smagorinsky_module = Some(self.kernel_manager.get_module("smagorinsky_kernels")?);
+        self.smagorinsky_module = Some(self.kernel_manager.get_module("smagorinsky_kernels")?.clone());
         Ok(())
     }
 
@@ -138,20 +138,22 @@ impl SmagorinskyGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&mut buffers.delta)
-                .arg(&dx)
-                .arg(&dy)
-                .arg(&dz)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Filter width kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&mut buffers.delta)
+                    .arg(&dx)
+                    .arg(&dy)
+                    .arg(&dz)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Filter width kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -216,31 +218,33 @@ impl SmagorinskyGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(u)
-                .arg(v)
-                .arg(w)
-                .arg(&mut buffers.dudx)
-                .arg(&mut buffers.dudy)
-                .arg(&mut buffers.dudz)
-                .arg(&mut buffers.dvdx)
-                .arg(&mut buffers.dvdy)
-                .arg(&mut buffers.dvdz)
-                .arg(&mut buffers.dwdx)
-                .arg(&mut buffers.dwdy)
-                .arg(&mut buffers.dwdz)
-                .arg(&dx)
-                .arg(&dy)
-                .arg(&dz)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Velocity gradients kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(u)
+                    .arg(v)
+                    .arg(w)
+                    .arg(&mut buffers.dudx)
+                    .arg(&mut buffers.dudy)
+                    .arg(&mut buffers.dudz)
+                    .arg(&mut buffers.dvdx)
+                    .arg(&mut buffers.dvdy)
+                    .arg(&mut buffers.dvdz)
+                    .arg(&mut buffers.dwdx)
+                    .arg(&mut buffers.dwdy)
+                    .arg(&mut buffers.dwdz)
+                    .arg(&dx)
+                    .arg(&dy)
+                    .arg(&dz)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Velocity gradients kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         Ok(())
@@ -274,26 +278,28 @@ impl SmagorinskyGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.dudx)
-                .arg(&buffers.dudy)
-                .arg(&buffers.dudz)
-                .arg(&buffers.dvdx)
-                .arg(&buffers.dvdy)
-                .arg(&buffers.dvdz)
-                .arg(&buffers.dwdx)
-                .arg(&buffers.dwdy)
-                .arg(&buffers.dwdz)
-                .arg(&mut buffers.strain_rate)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("Strain rate kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.dudx)
+                    .arg(&buffers.dudy)
+                    .arg(&buffers.dudz)
+                    .arg(&buffers.dvdx)
+                    .arg(&buffers.dvdy)
+                    .arg(&buffers.dvdz)
+                    .arg(&buffers.dwdx)
+                    .arg(&buffers.dwdy)
+                    .arg(&buffers.dwdz)
+                    .arg(&mut buffers.strain_rate)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("Strain rate kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         Ok(())
@@ -325,20 +331,22 @@ impl SmagorinskyGpuModel {
         };
 
         unsafe {
-            self.kernel_manager
-                .stream()
-                .launch_builder(&func)
-                .arg(&buffers.strain_rate)
-                .arg(&buffers.delta)
-                .arg(&mut buffers.sgs_viscosity)
-                .arg(&self.constants.cs)
-                .arg(&(buffers.nx as i32))
-                .arg(&(buffers.ny as i32))
-                .arg(&(buffers.nz as i32))
-                .launch(config)
-                .map_err(|e| {
-                    CfdError::gpu_error(&format!("SGS viscosity kernel launch failed: {}", e))
-                })?;
+            unsafe {
+                self.kernel_manager
+                    .stream()
+                    .launch_builder(&func)
+                    .arg(&buffers.strain_rate)
+                    .arg(&buffers.delta)
+                    .arg(&mut buffers.sgs_viscosity)
+                    .arg(&self.constants.cs)
+                    .arg(&(buffers.nx as i32))
+                    .arg(&(buffers.ny as i32))
+                    .arg(&(buffers.nz as i32))
+                    .launch(config)
+                    .map_err(|e| {
+                        CfdError::gpu_error(&format!("SGS viscosity kernel launch failed: {}", e))
+                    })?;
+            }
         }
 
         self.kernel_manager.synchronize()?;
@@ -356,7 +364,7 @@ impl SmagorinskyGpuModel {
         let mut sgs_viscosity_host = vec![0.0f32; n];
 
         self.kernel_manager
-            .stream
+            .stream()
             .memcpy_dtoh(&buffers.sgs_viscosity, &mut sgs_viscosity_host)
             .map_err(|e| {
                 CfdError::gpu_error(&format!("Failed to copy SGS viscosity from GPU: {}", e))