pass env data through grid2

docs/src/progress/renderertodo.md

@@ -1288,6 +1288,14 @@ Fix density advection dx being wrong
 Update gravity const
 Add bounds setting to velocity projection
 
+(12/11/2024)
+Fix multigrid solver
+Fix conjugate gradient solver
+grid2 starting to look realtime!
+
+(12/12/2024)
+Pass environment data through grid2 solver
+
 
 
 # TODO

src/main/c/includes/fluid/env/environment.h

@@ -81,6 +81,30 @@ typedef struct {
     double velocityProject;
 } FluidTimeTracking;
 
+/**
+ * The state for the grid2 simulator
+ */
+typedef struct {
+    /**
+     * A grid that stores a mask of the border locations
+     */
+    float * fluid_grid2_border_mask;
+    float * fluid_grid2_border_mask_inverted;
+
+    /**
+     * An array that stores values for the neighbor's arrays
+     */
+    float * fluid_grid2_neighborArr_d;
+    float * fluid_grid2_neighborArr_d0;
+    float * fluid_grid2_neighborArr_u;
+    float * fluid_grid2_neighborArr_v;
+    float * fluid_grid2_neighborArr_w;
+    float * fluid_grid2_neighborArr_u0;
+    float * fluid_grid2_neighborArr_v0;
+    float * fluid_grid2_neighborArr_w0;
+    float * fluid_grid2_neighborArr_bounds;
+} FluidGrid2State;
+
 /**
  * Used for tracking the change in density over time of the environment
  */
@@ -90,6 +114,7 @@ typedef struct {
     double normalizationRatio;
     int frame;
     FluidTimeTracking timeTracking;
+    FluidGrid2State grid2;
 } FluidSimState;
 
 /**

src/main/c/includes/fluid/sim/grid2/density.h

@@ -10,7 +10,7 @@
 /**
  * Actually performs the advection
 */
-void fluid_grid2_advect(uint32_t chunk_mask, int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt);
+void fluid_grid2_advect(Environment * environment, uint32_t chunk_mask, int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt);
 
 
 
@@ -32,9 +32,10 @@ void fluid_grid2_addDensity(
  * A single iteration of the jacobi to solve density diffusion
  */
 LIBRARY_API void fluid_grid2_solveDiffuseDensity(
-  float ** d,
+    Environment * environment,
-  float ** d0,
+    float ** d,
-  float dt
+    float ** d0,
+    float dt
 );
 
 
@@ -44,14 +45,14 @@ LIBRARY_API void fluid_grid2_solveDiffuseDensity(
 /**
  * Advects the density based on the vectors
 */
-LIBRARY_API void fluid_grid2_advectDensity(float ** d, float ** d0, float ** ur, float ** vr, float ** wr, float dt);
+LIBRARY_API void fluid_grid2_advectDensity(Environment * environment, float ** d, float ** d0, float ** ur, float ** vr, float ** wr, float dt);
 
 
 
 /**
  * Normalizes the density array with a given ratio
  */
-void fluid_grid2_normalizeDensity(float ** d, float ratio);
+void fluid_grid2_normalizeDensity(Environment * environment, float ** d, float ratio);
 
 
 

src/main/c/includes/fluid/sim/grid2/grid2.h

@@ -31,19 +31,6 @@ LIBRARY_API void fluid_grid2_simulate(int numChunks, Chunk ** passedInChunks, En
 
 
 
-/**
- * Allocates the arrays necessary for grid2 simulation
- */
-LIBRARY_API void fluid_grid2_allocate_arrays();
-
-
-
-
-
-
-
-
-
 
 
 

src/main/c/includes/fluid/sim/grid2/utilities.h

@@ -4,6 +4,8 @@
 
 #include <stdint.h>
 
+#include "fluid/env/environment.h"
+
 
 
 /**
@@ -17,8 +19,9 @@ void fluid_grid2_add_source(float * x, float * s, float dt);
  * Sets the bounds of this cube to those of its neighbor
 */
 LIBRARY_API void fluid_grid2_set_bounds(
-  int vector_dir,
+    Environment * environment,
-  float * target
+    int vector_dir,
+    float * target
 );
 
 

src/main/c/includes/fluid/sim/grid2/velocity.h

@@ -33,6 +33,7 @@
  * Adds the sources to the destinations
  */
 void fluid_grid2_addSourceToVectors(
+  Environment * environment,
   float ** jru,
   float ** jrv,
   float ** jrw,
@@ -57,6 +58,7 @@ void fluid_grid2_addSourceToVectors(
  * @param dt The timestep for the simulation
  */
 LIBRARY_API void fluid_grid2_setupProjection(
+  Environment * environment,
   float ** ur,
   float ** vr,
   float ** wr,
@@ -74,6 +76,7 @@ LIBRARY_API void fluid_grid2_setupProjection(
  * @param jrv0 The first derivative field
  */
 LIBRARY_API void fluid_grid2_solveProjection(
+  Environment * environment,
   Chunk * chunk,
   float ** jru0,
   float ** jrv0,
@@ -87,6 +90,7 @@ LIBRARY_API void fluid_grid2_solveProjection(
  * Thus we are left with an approximately mass-conserved field. 
  */
 LIBRARY_API void fluid_grid2_finalizeProjection(
+  Environment * environment,
   float ** jru,
   float ** jrv,
   float ** jrw,
@@ -101,6 +105,7 @@ LIBRARY_API void fluid_grid2_finalizeProjection(
  * Advects u, v, and w
  */
 LIBRARY_API void fluid_grid2_advectVectors(
+  Environment * environment,
   float ** jru,
   float ** jrv,
   float ** jrw,
@@ -113,13 +118,14 @@ LIBRARY_API void fluid_grid2_advectVectors(
 /**
  * Actually performs the advection
 */
-void fluid_grid2_advect_velocity(int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt);
+void fluid_grid2_advect_velocity(Environment * environment, int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt);
 
 
 /*
  * Solves vector diffusion along all axis
  */
 LIBRARY_API void fluid_grid2_solveVectorDiffuse (
+  Environment * environment,
   float ** jru,
   float ** jrv,
   float ** jrw,

src/main/c/src/fluid/env/environment.c

@@ -2,8 +2,13 @@
 
 #include "public.h"
 #include "fluid/env/environment.h"
+#include "fluid/env/utilities.h"
 
 
+
+
+void fluid_environment_allocate_arrays(Environment * environment);
+
 /**
  * Creates an environment
  */
@@ -12,6 +17,10 @@ LIBRARY_API Environment * fluid_environment_create(){
     rVal->queue.cellularQueue = NULL;
     rVal->queue.gridQueue = NULL;
     rVal->queue.grid2Queue = NULL;
+
+    //allocate arrays
+    fluid_environment_allocate_arrays(rVal);
+
     return rVal;
 }
 
@@ -20,4 +29,36 @@ LIBRARY_API Environment * fluid_environment_create(){
  */
 LIBRARY_API void fluid_environment_free(Environment * environment){
     free(environment);
+}
+
+
+
+/**
+ * Allocates the arrays necessary for environment simulation
+ */
+void fluid_environment_allocate_arrays(Environment * environment){
+    environment->state.grid2.fluid_grid2_border_mask = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_border_mask_inverted = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    for(int x = 0; x < DIM; x++){
+        for(int y = 0; y < DIM; y++){
+            for(int z = 0; z < DIM; z++){
+                if(x == 0 || x == DIM-1 || y == 0 || y == DIM-1 || z == 0 || z == DIM-1){
+                    environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)] = 1;
+                    environment->state.grid2.fluid_grid2_border_mask_inverted[IX(x,y,z)] = 0;
+                } else {
+                    environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)] = 0;
+                    environment->state.grid2.fluid_grid2_border_mask_inverted[IX(x,y,z)] = 1;
+                }
+            }
+        }
+    }
+    environment->state.grid2.fluid_grid2_neighborArr_d = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_d0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_u = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_v = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_w = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_u0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_v0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_w0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
+    environment->state.grid2.fluid_grid2_neighborArr_bounds = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
 }

src/main/c/src/fluid/queue/javainterface.c

@@ -140,10 +140,6 @@ JNIEXPORT void JNICALL Java_electrosphere_server_fluid_simulator_FluidAccelerate
     environment->lookupTable.serverFluidChunkTable.asleepId = (*env)->GetFieldID(env,fluidSimStorageClass,"asleep","Z");
     environment->lookupTable.serverFluidChunkTable.homogenousId = (*env)->GetFieldID(env,fluidSimStorageClass,"isHomogenous","Z");
     environment->lookupTable.serverFluidChunkTable.normalizationRatioId = (*env)->GetStaticFieldID(env,fluidSimStorageClass,"normalizationRatio","F");
-
-
-    //init grid2 sim
-    fluid_grid2_allocate_arrays();
 }
 
 /**

src/main/c/src/fluid/sim/grid2/density.c

@@ -58,9 +58,10 @@ void fluid_grid2_addDensity(
  * A single iteration of the jacobi to solve density diffusion
  */
 LIBRARY_API void fluid_grid2_solveDiffuseDensity(
-  float ** d,
+    Environment * environment,
-  float ** d0,
+    float ** d,
-  float dt
+    float ** d0,
+    float dt
 ){
     float a=dt*FLUID_GRID2_DIFFUSION_CONSTANT/(FLUID_GRID2_H*FLUID_GRID2_H);
     float c=1+6*a;
@@ -81,14 +82,14 @@ LIBRARY_API void fluid_grid2_solveDiffuseDensity(
         solver_gauss_seidel_iterate_parallel(x,x0,a,c,DIM);
 
         //set bounds
-        fluid_grid2_set_bounds(0,x);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,x);
     }
 }
 
 /**
  * Advects the density based on the vectors
 */
-LIBRARY_API void fluid_grid2_advectDensity(float ** d, float ** d0, float ** ur, float ** vr, float ** wr, float dt){
+LIBRARY_API void fluid_grid2_advectDensity(Environment * environment, float ** d, float ** d0, float ** ur, float ** vr, float ** wr, float dt){
     int i, j, k, i0, j0, k0, i1, j1, k1;
     int m,n,o;
     float x, y, z, s0, t0, s1, t1, u1, u0, dtx,dty,dtz;
@@ -235,13 +236,13 @@ LIBRARY_API void fluid_grid2_advectDensity(float ** d, float ** d0, float ** ur,
         }
     }
     //set bounds
-    fluid_grid2_set_bounds(0,center_d);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,center_d);
 }
 
 /**
  * Normalizes the density array with a given ratio
  */
-void fluid_grid2_normalizeDensity(float ** d, float ratio){
+void fluid_grid2_normalizeDensity(Environment * environment, float ** d, float ratio){
     int j;
     int size=DIM*DIM*DIM;
     float * x = GET_ARR_RAW(d,CENTER_LOC);

src/main/c/src/fluid/sim/grid2/grid2.c

@@ -23,26 +23,7 @@
 static inline void fluid_grid2_saveStep(float * values, const char * name);
 static inline void fluid_grid2_applyGravity(Chunk * currentChunk, Environment * environment);
 static inline void fluid_grid2_clearArr(float ** d);
-static inline void fluid_grid2_rewrite_bounds(Chunk * chunk);
+static inline void fluid_grid2_rewrite_bounds(Environment * environment, Chunk * chunk);
-
-/**
- * A grid that stores a mask of the border locations
- */
-float * fluid_grid2_border_mask;
-float * fluid_grid2_border_mask_inverted;
-
-/**
- * An array that stores values for the neighbor's arrays
- */
-float * fluid_grid2_neighborArr_d;
-float * fluid_grid2_neighborArr_d0;
-float * fluid_grid2_neighborArr_u;
-float * fluid_grid2_neighborArr_v;
-float * fluid_grid2_neighborArr_w;
-float * fluid_grid2_neighborArr_u0;
-float * fluid_grid2_neighborArr_v0;
-float * fluid_grid2_neighborArr_w0;
-float * fluid_grid2_neighborArr_bounds;
 
 /**
  * Used for storing timings
@@ -66,15 +47,15 @@ LIBRARY_API void fluid_grid2_simulate(
     //
     for(int i = 0; i < numChunks; i++){
         Chunk * currentChunk = chunks[i];
-
         //update the bounds arrays
-        fluid_grid2_rewrite_bounds(currentChunk);
+        fluid_grid2_rewrite_bounds(environment,currentChunk);
 
 
 
         //add velocity
         fluid_grid2_applyGravity(currentChunk,environment);
         fluid_grid2_addSourceToVectors(
+            environment,
             currentChunk->u,
             currentChunk->v,
             currentChunk->w,
@@ -90,7 +71,7 @@ LIBRARY_API void fluid_grid2_simulate(
         fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
 
         //solve vector diffusion
-        fluid_grid2_solveVectorDiffuse(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,timestep);
+        fluid_grid2_solveVectorDiffuse(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,timestep);
     // }
 
     // //time tracking
@@ -102,15 +83,17 @@ LIBRARY_API void fluid_grid2_simulate(
 
     // for(int i = 0; i < numChunks; i++){
     //     Chunk * currentChunk = chunks[i];
-        //setup projection
+    //     //update the bounds arrays
-        fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
+    //     fluid_grid2_rewrite_bounds(currentChunk);
+        // setup projection
+        fluid_grid2_setupProjection(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
 
         //
         //Perform main projection solver
-        fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,timestep);
+        fluid_grid2_solveProjection(environment,currentChunk,currentChunk->u0,currentChunk->v0,timestep);
 
         //Finalize projection
-        fluid_grid2_finalizeProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
+        fluid_grid2_finalizeProjection(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
 
         //swap all vector fields
         fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
@@ -127,8 +110,11 @@ LIBRARY_API void fluid_grid2_simulate(
 
     // for(int i = 0; i < numChunks; i++){
     //     Chunk * currentChunk = chunks[i];
-        //advect
+    //     Chunk * currentChunk = chunks[i];
-        fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,timestep);
+    //     //update the bounds arrays
+    //     fluid_grid2_rewrite_bounds(currentChunk);
+        // advect
+        fluid_grid2_advectVectors(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,timestep);
 
     // }
 
@@ -141,12 +127,14 @@ LIBRARY_API void fluid_grid2_simulate(
 
     // for(int i = 0; i < numChunks; i++){
     //     Chunk * currentChunk = chunks[i];
+    //     //update the bounds arrays
+    //     fluid_grid2_rewrite_bounds(currentChunk);
         //setup projection
-        fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
+        fluid_grid2_setupProjection(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
         //Perform main projection solver
-        fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,timestep);
+        fluid_grid2_solveProjection(environment,currentChunk,currentChunk->u0,currentChunk->v0,timestep);
         //Finalize projection
-        fluid_grid2_finalizeProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
+        fluid_grid2_finalizeProjection(environment,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,timestep);
     }
 
 
@@ -174,13 +162,15 @@ LIBRARY_API void fluid_grid2_simulate(
     environment->state.newDensity = 0;
     for(int i = 0; i < numChunks; i++){
         Chunk * currentChunk = chunks[i];
+        //update the bounds arrays
+        fluid_grid2_rewrite_bounds(environment, currentChunk);
 
         fluid_grid2_addDensity(environment,currentChunk->d,currentChunk->d0,timestep);
         environment->state.existingDensity = environment->state.existingDensity + fluid_grid2_calculateSum(currentChunk->d);
         //swap all density arrays
         fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
         //diffuse density
-        fluid_grid2_solveDiffuseDensity(currentChunk->d,currentChunk->d0,timestep);
+        fluid_grid2_solveDiffuseDensity(environment,currentChunk->d,currentChunk->d0,timestep);
         //swap all density arrays
         fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
     }
@@ -194,8 +184,10 @@ LIBRARY_API void fluid_grid2_simulate(
     
     for(int i = 0; i < numChunks; i++){
         Chunk * currentChunk = chunks[i];
+        //update the bounds arrays
+        fluid_grid2_rewrite_bounds(environment, currentChunk);
         //advect density
-        fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,timestep);
+        fluid_grid2_advectDensity(environment,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,timestep);
     }
 
     //get time at end
@@ -220,7 +212,9 @@ LIBRARY_API void fluid_grid2_simulate(
     {
         for(int i = 0; i < numChunks; i++){
             Chunk * currentChunk = chunks[i];
-            fluid_grid2_set_bounds(0,currentChunk->d[CENTER_LOC]);
+            //update the bounds arrays
+            fluid_grid2_rewrite_bounds(environment, currentChunk);
+            fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,currentChunk->d[CENTER_LOC]);
         }
     }
 
@@ -242,7 +236,7 @@ LIBRARY_API void fluid_grid2_simulate(
         }
         for(int i = 0; i < numChunks; i++){
             Chunk * currentChunk = chunks[i];
-            fluid_grid2_normalizeDensity(currentChunk->d,normalizationRatio);
+            fluid_grid2_normalizeDensity(environment,currentChunk->d,normalizationRatio);
         }
     }
 
@@ -330,41 +324,10 @@ static inline void fluid_grid2_clearArr(float ** d){
     }
 }
 
-
-/**
- * Allocates the arrays necessary for grid2 simulation
- */
-LIBRARY_API void fluid_grid2_allocate_arrays(){
-    fluid_grid2_border_mask = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_border_mask_inverted = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    for(int x = 0; x < DIM; x++){
-        for(int y = 0; y < DIM; y++){
-            for(int z = 0; z < DIM; z++){
-                if(x == 0 || x == DIM-1 || y == 0 || y == DIM-1 || z == 0 || z == DIM-1){
-                    fluid_grid2_border_mask[IX(x,y,z)] = 1;
-                    fluid_grid2_border_mask_inverted[IX(x,y,z)] = 0;
-                } else {
-                    fluid_grid2_border_mask[IX(x,y,z)] = 0;
-                    fluid_grid2_border_mask_inverted[IX(x,y,z)] = 1;
-                }
-            }
-        }
-    }
-    fluid_grid2_neighborArr_d = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_d0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_u = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_v = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_w = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_u0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_v0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_w0 = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-    fluid_grid2_neighborArr_bounds = (float *)calloc(1,DIM * DIM * DIM * sizeof(float));
-}
-
 /**
  * Applies a bounds array to a source array
  */
-static inline void fluid_grid2_apply_bounds_mask(float * realArr, float * boundsArr){
+static inline void fluid_grid2_apply_bounds_mask(Environment * environment, float * realArr, float * boundsArr){
     __m256 maskedBounds, realVal, invertedMask, realPart, finalVec;
     int x;
     for(int z = 0; z < 18; z++){
@@ -372,10 +335,10 @@ static inline void fluid_grid2_apply_bounds_mask(float * realArr, float * bounds
             //lower part
             x = 0;
             //border part
-            maskedBounds = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskedBounds = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             //real part
             realVal = _mm256_loadu_ps(&realArr[IX(x,y,z)]);
-            invertedMask = _mm256_loadu_ps(&fluid_grid2_border_mask_inverted[IX(x,y,z)]);
+            invertedMask = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask_inverted[IX(x,y,z)]);
             realPart = _mm256_mul_ps(realVal,invertedMask);
             finalVec = _mm256_add_ps(realPart,maskedBounds);
             _mm256_storeu_ps(&realArr[IX(x,y,z)],finalVec);
@@ -384,10 +347,10 @@ static inline void fluid_grid2_apply_bounds_mask(float * realArr, float * bounds
             //middle part
             x = 8;
             //border part
-            maskedBounds = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskedBounds = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             //real part
             realVal = _mm256_loadu_ps(&realArr[IX(x,y,z)]);
-            invertedMask = _mm256_loadu_ps(&fluid_grid2_border_mask_inverted[IX(x,y,z)]);
+            invertedMask = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask_inverted[IX(x,y,z)]);
             realPart = _mm256_mul_ps(realVal,invertedMask);
             finalVec = _mm256_add_ps(realPart,maskedBounds);
             _mm256_storeu_ps(&realArr[IX(x,y,z)],finalVec);
@@ -395,10 +358,10 @@ static inline void fluid_grid2_apply_bounds_mask(float * realArr, float * bounds
             //upper part
             x = 10;
             //border part
-            maskedBounds = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskedBounds = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             //real part
             realVal = _mm256_loadu_ps(&realArr[IX(x,y,z)]);
-            invertedMask = _mm256_loadu_ps(&fluid_grid2_border_mask_inverted[IX(x,y,z)]);
+            invertedMask = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask_inverted[IX(x,y,z)]);
             realPart = _mm256_mul_ps(realVal,invertedMask);
             finalVec = _mm256_add_ps(realPart,maskedBounds);
             _mm256_storeu_ps(&realArr[IX(x,y,z)],finalVec);
@@ -410,23 +373,23 @@ static inline void fluid_grid2_apply_bounds_mask(float * realArr, float * bounds
 /**
  * Applies the bounds mask to the current chunk
  */
-static inline void fluid_grid2_apply_neighbors(Chunk * chunk){
+static inline void fluid_grid2_apply_neighbors(Environment * environment, Chunk * chunk){
-    fluid_grid2_apply_bounds_mask(chunk->d[CENTER_LOC], fluid_grid2_neighborArr_d);
+    fluid_grid2_apply_bounds_mask(environment,chunk->d[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_d);
-    fluid_grid2_apply_bounds_mask(chunk->d0[CENTER_LOC], fluid_grid2_neighborArr_d0);
+    fluid_grid2_apply_bounds_mask(environment,chunk->d0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_d0);
-    fluid_grid2_apply_bounds_mask(chunk->u[CENTER_LOC], fluid_grid2_neighborArr_u);
+    fluid_grid2_apply_bounds_mask(environment,chunk->u[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_u);
-    fluid_grid2_apply_bounds_mask(chunk->v[CENTER_LOC], fluid_grid2_neighborArr_v);
+    fluid_grid2_apply_bounds_mask(environment,chunk->v[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_v);
-    fluid_grid2_apply_bounds_mask(chunk->w[CENTER_LOC], fluid_grid2_neighborArr_w);
+    fluid_grid2_apply_bounds_mask(environment,chunk->w[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_w);
-    fluid_grid2_apply_bounds_mask(chunk->u0[CENTER_LOC], fluid_grid2_neighborArr_u0);
+    fluid_grid2_apply_bounds_mask(environment,chunk->u0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_u0);
-    fluid_grid2_apply_bounds_mask(chunk->v0[CENTER_LOC], fluid_grid2_neighborArr_v0);
+    fluid_grid2_apply_bounds_mask(environment,chunk->v0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_v0);
-    fluid_grid2_apply_bounds_mask(chunk->w0[CENTER_LOC], fluid_grid2_neighborArr_w0);
+    fluid_grid2_apply_bounds_mask(environment,chunk->w0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_w0);
-    fluid_grid2_apply_bounds_mask(chunk->bounds[CENTER_LOC], fluid_grid2_neighborArr_bounds);
+    fluid_grid2_apply_bounds_mask(environment,chunk->bounds[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_bounds);
 }
 
 
 /**
  * Quickly masks a chunk's arrays
  */
-static inline void fluid_grid2_populate_masked_arr(float * sourceArr, float * workingArr){
+static inline void fluid_grid2_populate_masked_arr(Environment * environment, float * sourceArr, float * workingArr){
     __m256 arrVal, maskVal, masked;
     int x;
     for(int z = 0; z < 18; z++){
@@ -434,21 +397,21 @@ static inline void fluid_grid2_populate_masked_arr(float * sourceArr, float * wo
             //lower part
             x = 0;
             arrVal = _mm256_loadu_ps(&sourceArr[IX(x,y,z)]);
-            maskVal = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskVal = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             masked = _mm256_mul_ps(arrVal,maskVal);
             _mm256_storeu_ps(&workingArr[IX(x,y,z)],masked);
 
             //middle part
             x = 8;
             arrVal = _mm256_loadu_ps(&sourceArr[IX(x,y,z)]);
-            maskVal = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskVal = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             masked = _mm256_mul_ps(arrVal,maskVal);
             _mm256_storeu_ps(&workingArr[IX(x,y,z)],masked);
 
             //upper part
             x = 10;
             arrVal = _mm256_loadu_ps(&sourceArr[IX(x,y,z)]);
-            maskVal = _mm256_loadu_ps(&fluid_grid2_border_mask[IX(x,y,z)]);
+            maskVal = _mm256_loadu_ps(&environment->state.grid2.fluid_grid2_border_mask[IX(x,y,z)]);
             masked = _mm256_mul_ps(arrVal,maskVal);
             _mm256_storeu_ps(&workingArr[IX(x,y,z)],masked);
         }
@@ -458,16 +421,16 @@ static inline void fluid_grid2_populate_masked_arr(float * sourceArr, float * wo
 /**
  * Rewrites the bounds arrays to contain the bounds of the current chunk
  */
-static inline void fluid_grid2_rewrite_bounds(Chunk * chunk){
+static inline void fluid_grid2_rewrite_bounds(Environment * environment, Chunk * chunk){
-    fluid_grid2_populate_masked_arr(chunk->d[CENTER_LOC], fluid_grid2_neighborArr_d);
+    fluid_grid2_populate_masked_arr(environment,chunk->d[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_d);
-    fluid_grid2_populate_masked_arr(chunk->d0[CENTER_LOC], fluid_grid2_neighborArr_d0);
+    fluid_grid2_populate_masked_arr(environment,chunk->d0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_d0);
-    fluid_grid2_populate_masked_arr(chunk->u[CENTER_LOC], fluid_grid2_neighborArr_u);
+    fluid_grid2_populate_masked_arr(environment,chunk->u[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_u);
-    fluid_grid2_populate_masked_arr(chunk->v[CENTER_LOC], fluid_grid2_neighborArr_v);
+    fluid_grid2_populate_masked_arr(environment,chunk->v[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_v);
-    fluid_grid2_populate_masked_arr(chunk->w[CENTER_LOC], fluid_grid2_neighborArr_w);
+    fluid_grid2_populate_masked_arr(environment,chunk->w[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_w);
-    fluid_grid2_populate_masked_arr(chunk->u0[CENTER_LOC], fluid_grid2_neighborArr_u0);
+    fluid_grid2_populate_masked_arr(environment,chunk->u0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_u0);
-    fluid_grid2_populate_masked_arr(chunk->v0[CENTER_LOC], fluid_grid2_neighborArr_v0);
+    fluid_grid2_populate_masked_arr(environment,chunk->v0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_v0);
-    fluid_grid2_populate_masked_arr(chunk->w0[CENTER_LOC], fluid_grid2_neighborArr_w0);
+    fluid_grid2_populate_masked_arr(environment,chunk->w0[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_w0);
-    fluid_grid2_populate_masked_arr(chunk->bounds[CENTER_LOC], fluid_grid2_neighborArr_bounds);
+    fluid_grid2_populate_masked_arr(environment,chunk->bounds[CENTER_LOC], environment->state.grid2.fluid_grid2_neighborArr_bounds);
 }
 
 

src/main/c/src/fluid/sim/grid2/utilities.c

@@ -45,8 +45,9 @@ void fluid_grid2_add_source(float * x, float * s, float dt){
  * Sets the bounds of this cube to those of its neighbor
 */
 LIBRARY_API void fluid_grid2_set_bounds(
-  int vector_dir,
+    Environment * environment,
-  float * target
+    int vector_dir,
+    float * target
 ){
     //set the faces bounds
     for(int x=1; x < DIM-1; x++){

src/main/c/src/fluid/sim/grid2/velocity.c

@@ -15,22 +15,20 @@
 #define SET_BOUND_IGNORE 0
 #define SET_BOUND_USE_NEIGHBOR 1
 
-void fluid_grid2_add_source(float * x, float * s, float dt);
-void fluid_grid2_advect(uint32_t chunk_mask, int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt);
-
 
 /*
  * Adds force to all vectors
  */
-void fluid_grid2_addSourceToVectors
+void fluid_grid2_addSourceToVectors(
-  (
+    Environment * environment,
-  float ** jru,
+    float ** jru,
-  float ** jrv,
+    float ** jrv,
-  float ** jrw,
+    float ** jrw,
-  float ** jru0,
+    float ** jru0,
-  float ** jrv0,
+    float ** jrv0,
-  float ** jrw0,
+    float ** jrw0,
-  float dt){
+    float dt
+){
     fluid_grid2_add_source(GET_ARR_RAW(jru,CENTER_LOC),GET_ARR_RAW(jru0,CENTER_LOC),dt);
     fluid_grid2_add_source(GET_ARR_RAW(jrv,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),dt);
     fluid_grid2_add_source(GET_ARR_RAW(jrw,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
@@ -39,14 +37,15 @@ void fluid_grid2_addSourceToVectors
 /*
  * Solves vector diffusion along all axis
  */
-LIBRARY_API void fluid_grid2_solveVectorDiffuse (
+LIBRARY_API void fluid_grid2_solveVectorDiffuse(
-  float ** jru,
+    Environment * environment,
-  float ** jrv,
+    float ** jru,
-  float ** jrw,
+    float ** jrv,
-  float ** jru0,
+    float ** jrw,
-  float ** jrv0,
+    float ** jru0,
-  float ** jrw0,
+    float ** jrv0,
-  float dt
+    float ** jrw0,
+    float dt
 ){
     float a = dt*FLUID_GRID2_VISCOSITY_CONSTANT/(FLUID_GRID2_H*FLUID_GRID2_H);
     float c = 1+6*a;
@@ -69,9 +68,9 @@ LIBRARY_API void fluid_grid2_solveVectorDiffuse (
         solver_gauss_seidel_iterate_parallel(w,w0,a,c,DIM);
 
         //set bounds
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,u);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_U,u);
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,v);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_V,v);
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_W,w);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_W,w);
     }
 }
 
@@ -89,12 +88,13 @@ LIBRARY_API void fluid_grid2_solveVectorDiffuse (
  * @param dt The timestep for the simulation
  */
 LIBRARY_API void fluid_grid2_setupProjection(
-  float ** ur,
+    Environment * environment,
-  float ** vr,
+    float ** ur,
-  float ** wr,
+    float ** vr,
-  float ** pr,
+    float ** wr,
-  float ** divr,
+    float ** pr,
-  float dt
+    float ** divr,
+    float dt
 ){
     int i, j, k;
 
@@ -153,8 +153,8 @@ LIBRARY_API void fluid_grid2_setupProjection(
             
         }
     }
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,p);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,p);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,div);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,div);
 }
 
 /**
@@ -164,6 +164,7 @@ LIBRARY_API void fluid_grid2_setupProjection(
  * @param jrv0 The first derivative field
  */
 LIBRARY_API void fluid_grid2_solveProjection(
+    Environment * environment,
     Chunk * chunk,
     float ** jru0,
     float ** jrv0,
@@ -183,7 +184,7 @@ LIBRARY_API void fluid_grid2_solveProjection(
     chunk->projectionIterations = 0;
     while(chunk->projectionIterations < FLUID_GRID2_SOLVER_MULTIGRID_MAX_ITERATIONS && (chunk->projectionResidual > FLUID_GRID2_SOLVER_MULTIGRID_TOLERANCE || chunk->projectionResidual < -FLUID_GRID2_SOLVER_MULTIGRID_TOLERANCE)){
         chunk->projectionResidual = solver_multigrid_parallel_iterate(p,div,a,c);
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,p);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,p);
         chunk->projectionIterations++;
     }
     // if(chunk->projectionResidual > FLUID_GRID2_SOLVER_MULTIGRID_TOLERANCE || chunk->projectionResidual < -FLUID_GRID2_SOLVER_MULTIGRID_TOLERANCE){
@@ -201,7 +202,7 @@ LIBRARY_API void fluid_grid2_solveProjection(
     //solve with CG
     while(chunk->projectionIterations < FLUID_GRID2_SOLVER_CG_MAX_ITERATIONS && (chunk->projectionResidual > FLUID_GRID2_SOLVER_CG_TOLERANCE || chunk->projectionResidual < -FLUID_GRID2_SOLVER_CG_TOLERANCE)){
         chunk->projectionResidual = solver_conjugate_gradient_iterate_parallel(p,div,a,c);;
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,p);
+        fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_NO_DIR,p);
         chunk->projectionIterations++;
     }
     if(chunk->projectionResidual > FLUID_GRID2_SOLVER_CG_TOLERANCE || chunk->projectionResidual < -FLUID_GRID2_SOLVER_CG_TOLERANCE){
@@ -215,12 +216,13 @@ LIBRARY_API void fluid_grid2_solveProjection(
  * Thus we are left with an approximately mass-conserved field. 
  */
 LIBRARY_API void fluid_grid2_finalizeProjection(
-  float ** jru,
+    Environment * environment,
-  float ** jrv,
+    float ** jru,
-  float ** jrw,
+    float ** jrv,
-  float ** jru0,
+    float ** jrw,
-  float ** jrv0,
+    float ** jru0,
-  float dt
+    float ** jrv0,
+    float dt
 ){
     int i, j, k;
     __m256 constScalar = _mm256_set1_ps(2.0f*FLUID_GRID2_H);
@@ -289,32 +291,33 @@ LIBRARY_API void fluid_grid2_finalizeProjection(
         }
     }
     //set bounds
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,u);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_U,u);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,v);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_V,v);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_W,w);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_W,w);
 }
 
 /*
  * Advects u, v, and w
  */
 LIBRARY_API void fluid_grid2_advectVectors(
-  float ** jru,
+    Environment * environment,
-  float ** jrv,
+    float ** jru,
-  float ** jrw,
+    float ** jrv,
-  float ** jru0,
+    float ** jrw,
-  float ** jrv0,
+    float ** jru0,
-  float ** jrw0,
+    float ** jrv0,
-  float dt
+    float ** jrw0,
+    float dt
 ){
-    fluid_grid2_advect_velocity(1,jru,jru0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
+    fluid_grid2_advect_velocity(environment,FLUID_GRID2_BOUND_DIR_U,jru,jru0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
-    fluid_grid2_advect_velocity(2,jrv,jrv0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
+    fluid_grid2_advect_velocity(environment,FLUID_GRID2_BOUND_DIR_V,jrv,jrv0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
-    fluid_grid2_advect_velocity(3,jrw,jrw0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
+    fluid_grid2_advect_velocity(environment,FLUID_GRID2_BOUND_DIR_W,jrw,jrw0,GET_ARR_RAW(jru0,CENTER_LOC),GET_ARR_RAW(jrv0,CENTER_LOC),GET_ARR_RAW(jrw0,CENTER_LOC),dt);
 }
 
 /**
  * Actually performs the advection
 */
-void fluid_grid2_advect_velocity(int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt){
+void fluid_grid2_advect_velocity(Environment * environment, int b, float ** jrd, float ** jrd0, float * u, float * v, float * w, float dt){
     int i, j, k, i0, j0, k0, i1, j1, k1;
     int m,n,o;
     float x, y, z, s0, t0, s1, t1, u1, u0, dtx,dty,dtz;
@@ -461,7 +464,7 @@ void fluid_grid2_advect_velocity(int b, float ** jrd, float ** jrd0, float * u,
         }
     }
     //set bounds
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,u);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_U,u);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,v);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_V,v);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_W,w);
+    fluid_grid2_set_bounds(environment,FLUID_GRID2_BOUND_DIR_W,w);
 }

src/test/c/fluid/sim/grid2/add_dens_tests.c

@@ -74,8 +74,6 @@ int fluid_sim_grid2_add_dens_test1(){
 int fluid_sim_grid2_add_dens_tests(int argc, char **argv){
     int rVal = 0;
 
-    fluid_grid2_allocate_arrays();
-
     rVal += fluid_sim_grid2_add_dens_test1();
 
     return rVal;

src/test/c/fluid/sim/grid2/advect_projection_tests.c

@@ -53,22 +53,22 @@ int fluid_sim_grid2_advect_projection_test1(){
             fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
-            fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectVectors(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
 
             ////project
-            fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,currentChunk->u0[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_U,currentChunk->u0[CENTER_LOC]);
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,currentChunk->v0[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_V,currentChunk->v0[CENTER_LOC]);
-            fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_solveProjection(env,currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
-            fluid_grid2_finalizeProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_finalizeProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
             //advect density
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-            fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
         }
     }
 
@@ -118,22 +118,22 @@ int fluid_sim_grid2_advect_projection_compute_error_over_time(){
                     fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
                     fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
                     fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
-                    fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+                    fluid_grid2_advectVectors(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
                     fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
                     fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
                     fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
 
                     ////project
-                    fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+                    fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-                    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,currentChunk->u0[CENTER_LOC]);
+                    fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_U,currentChunk->u0[CENTER_LOC]);
-                    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,currentChunk->v0[CENTER_LOC]);
+                    fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_V,currentChunk->v0[CENTER_LOC]);
-                    fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+                    fluid_grid2_solveProjection(env,currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-                    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
+                    fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
-                    fluid_grid2_finalizeProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+                    fluid_grid2_finalizeProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
                     //advect density
                     fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-                    fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+                    fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
                 }
             }
             //test the result

src/test/c/fluid/sim/grid2/convergence_tests.c

@@ -163,8 +163,6 @@ int fluid_sim_grid2_convergence_test3(){
 int fluid_sim_grid2_convergence_tests(){
     int rVal = 0;
 
-    fluid_grid2_allocate_arrays();
-
     rVal += fluid_sim_grid2_convergence_test1();
     rVal += fluid_sim_grid2_convergence_test2();
     rVal += fluid_sim_grid2_convergence_test3();

src/test/c/fluid/sim/grid2/density_advection_tests.c

@@ -45,7 +45,7 @@ int fluid_sim_grid2_density_advection_test1(){
 
     //actually advect
     fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-    fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
 
     //sum the result
     float afterSum = chunk_queue_sum_density(queue);
@@ -82,7 +82,7 @@ int fluid_sim_grid2_density_advection_test2(){
         for(int chunkIndex = 0; chunkIndex < 1; chunkIndex++){
             currentChunk = queue[chunkIndex];
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-            fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
         }
     }
@@ -122,7 +122,7 @@ int fluid_sim_grid2_density_advection_test3(){
         for(int chunkIndex = 0; chunkIndex < 1; chunkIndex++){
             currentChunk = queue[chunkIndex];
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-            fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
         }
     }
@@ -165,7 +165,7 @@ int fluid_sim_grid2_density_advection_test4(){
         for(int chunkIndex = 0; chunkIndex < 1; chunkIndex++){
             currentChunk = queue[chunkIndex];
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-            fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
         }
     }
@@ -205,7 +205,7 @@ int fluid_sim_grid2_density_advection_test5(){
         for(int chunkIndex = 0; chunkIndex < 1; chunkIndex++){
             currentChunk = queue[chunkIndex];
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
-            fluid_grid2_advectDensity(currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectDensity(env,currentChunk->d,currentChunk->d0,currentChunk->u,currentChunk->v,currentChunk->w,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->d,currentChunk->d0);
         }
     }

src/test/c/fluid/sim/grid2/density_diffuse_tests.c

@@ -9,6 +9,7 @@
 #include "fluid/sim/grid2/density.h"
 #include "fluid/sim/grid2/solver_consts.h"
 #include "fluid/sim/grid2/utilities.h"
+#include "fluid/sim/grid2/velocity.h"
 #include "math/ode/multigrid.h"
 #include "../../../util/chunk_test_utils.h"
 #include "../../../util/test.h"
@@ -76,8 +77,8 @@ int fluid_sim_grid2_density_diffuse_test1(){
     }
     //diffuse density
     for(int l = 0; l < FLUID_GRID2_LINEARSOLVERTIMES; l++){
-            fluid_grid2_solveDiffuseDensity(currentChunk->d,currentChunk->d0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_solveDiffuseDensity(env,currentChunk->d,currentChunk->d0,FLUID_GRID2_SIM_STEP);
-            fluid_grid2_set_bounds(0,currentChunk->d[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_NO_DIR,currentChunk->d[CENTER_LOC]);
     }
     //swap all density arrays
     //swap vector fields
@@ -134,8 +135,8 @@ int fluid_sim_grid2_density_diffuse_test2(){
         }
         //diffuse density
         for(int l = 0; l < FLUID_GRID2_LINEARSOLVERTIMES; l++){
-                fluid_grid2_solveDiffuseDensity(currentChunk->d,currentChunk->d0,FLUID_GRID2_SIM_STEP);
+                fluid_grid2_solveDiffuseDensity(env,currentChunk->d,currentChunk->d0,FLUID_GRID2_SIM_STEP);
-                fluid_grid2_set_bounds(0,currentChunk->d[CENTER_LOC]);
+                fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_NO_DIR,currentChunk->d[CENTER_LOC]);
         }
         //swap all density arrays
         //swap vector fields

src/test/c/fluid/sim/grid2/finalize_projection_tests.c

@@ -35,12 +35,12 @@ int fluid_sim_grid2_finalize_projection_test1(){
     //setup chunk values
     Chunk * currentChunk = queue[0];
     currentChunk->u[CENTER_LOC][IX(3,3,3)] = 1.0f;
-    fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-    fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_solveProjection(env,currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
-    fluid_grid2_set_bounds(FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
+    fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_NO_DIR,currentChunk->u0[CENTER_LOC]);
 
     //finalize
-    fluid_grid2_finalizeProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_finalizeProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
     //test the result
     float expected, actual;

src/test/c/fluid/sim/grid2/full_sim_tests.c

@@ -137,8 +137,6 @@ int fluid_sim_grid2_full_sim_test3(){
 int fluid_sim_grid2_full_sim_tests(int argc, char **argv){
     int rVal = 0;
 
-    fluid_grid2_allocate_arrays();
-
     rVal += fluid_sim_grid2_full_sim_test1();
     rVal += fluid_sim_grid2_full_sim_test2();
     rVal += fluid_sim_grid2_full_sim_test3();

src/test/c/fluid/sim/grid2/setup_projection_tests.c

@@ -38,7 +38,7 @@ int fluid_sim_grid2_setup_projection_test1(){
     currentChunk->u[CENTER_LOC][IX(3,3,3)] = 1.0f;
 
     //actually simulate
-    fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
     //test the result
     rVal += assertEqualsFloat(currentChunk->v0[CENTER_LOC][IX(2,3,3)],-0.5f * FLUID_GRID2_H,"Divergence of the vector field at 3,3,3 should be -0.5 * h!   actual: %f    expected: %f  \n");
@@ -66,7 +66,7 @@ int fluid_sim_grid2_setup_projection_test2(){
     currentChunk->u[CENTER_LOC][IX(3,3,3)] = -1.0f;
 
     //actually simulate
-    fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
     //test the result
     rVal += assertEqualsFloat(currentChunk->v0[CENTER_LOC][IX(2,3,3)],0.5f * FLUID_GRID2_H,"Divergence of the vector field at 3,3,3 should be 0.5 * h!   actual: %f    expected: %f  \n");

src/test/c/fluid/sim/grid2/solve_projection_tests.c

@@ -35,10 +35,10 @@ int fluid_sim_grid2_solve_projection_test1(){
     //setup chunk values
     Chunk * currentChunk = queue[0];
     currentChunk->u[CENTER_LOC][IX(3,3,3)] = 1.0f;
-    fluid_grid2_setupProjection(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_setupProjection(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
     //actually solve
-    fluid_grid2_solveProjection(currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
+    fluid_grid2_solveProjection(env,currentChunk,currentChunk->u0,currentChunk->v0,FLUID_GRID2_SIM_STEP);
 
     //test the result
     float expected, actual;

src/test/c/fluid/sim/grid2/speed_tests.c

@@ -248,8 +248,6 @@ int fluid_sim_grid2_speed_test3(){
 int fluid_sim_grid2_speed_tests(){
     int rVal = 0;
 
-    fluid_grid2_allocate_arrays();
-
     rVal += fluid_sim_grid2_speed_test1();
     rVal += fluid_sim_grid2_speed_test2();
     rVal += fluid_sim_grid2_speed_test3();

src/test/c/fluid/sim/grid2/velocity_advection_tests.c

@@ -47,7 +47,7 @@ int fluid_sim_grid2_velocity_advection_test1(){
         fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
-            fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectVectors(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
@@ -100,7 +100,7 @@ int fluid_sim_grid2_velocity_advection_test2(){
             fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
-            fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_advectVectors(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
             fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
             fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
             fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
@@ -151,7 +151,7 @@ int fluid_sim_grid2_velocity_advection_test3(){
         fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
         fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
         fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);
-        fluid_grid2_advectVectors(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+        fluid_grid2_advectVectors(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
         fluid_grid2_flip_arrays(currentChunk->u,currentChunk->u0);
         fluid_grid2_flip_arrays(currentChunk->v,currentChunk->v0);
         fluid_grid2_flip_arrays(currentChunk->w,currentChunk->w0);

src/test/c/fluid/sim/grid2/velocity_diffuse_tests.c

@@ -55,11 +55,11 @@ int fluid_sim_grid2_velocity_diffuse_test1(){
     //solve vector diffusion
     for(int l = 0; l < FLUID_GRID2_LINEARSOLVERTIMES; l++){
         //solve vector diffusion
-        fluid_grid2_solveVectorDiffuse(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+        fluid_grid2_solveVectorDiffuse(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
         //update array for vectors
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,currentChunk->u[CENTER_LOC]);
+        fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_U,currentChunk->u[CENTER_LOC]);
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,currentChunk->v[CENTER_LOC]);
+        fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_V,currentChunk->v[CENTER_LOC]);
-        fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_W,currentChunk->w[CENTER_LOC]);
+        fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_W,currentChunk->w[CENTER_LOC]);
     }
     //swap all density arrays
     //swap vector fields
@@ -115,11 +115,11 @@ int fluid_sim_grid2_velocity_diffuse_test2(){
         //solve vector diffusion
         for(int l = 0; l < FLUID_GRID2_LINEARSOLVERTIMES; l++){
             //solve vector diffusion
-            fluid_grid2_solveVectorDiffuse(currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
+            fluid_grid2_solveVectorDiffuse(env,currentChunk->u,currentChunk->v,currentChunk->w,currentChunk->u0,currentChunk->v0,currentChunk->w0,FLUID_GRID2_SIM_STEP);
             //update array for vectors
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_U,currentChunk->u[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_U,currentChunk->u[CENTER_LOC]);
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_V,currentChunk->v[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_V,currentChunk->v[CENTER_LOC]);
-            fluid_grid2_set_bounds(FLUID_GRID2_BOUND_DIR_W,currentChunk->w[CENTER_LOC]);
+            fluid_grid2_set_bounds(env,FLUID_GRID2_BOUND_DIR_W,currentChunk->w[CENTER_LOC]);
         }
         //swap all density arrays
         //swap vector fields