diff --git a/include/flamegpu/gpu/CUDAAgentModel.h b/include/flamegpu/gpu/CUDAAgentModel.h
index c424edc8d..eb2fd943d 100644
--- a/include/flamegpu/gpu/CUDAAgentModel.h
+++ b/include/flamegpu/gpu/CUDAAgentModel.h
@@ -165,6 +165,10 @@ class CUDAAgentModel : public Simulation {
      * Number of times step() has been called since sim was last reset/init
      */
     unsigned int step_count;
+    /**
+     * Update the step counter for host and device.
+     */
+    void incrementStepCounter();
     /**
      * Map of agent storage 
      */
diff --git a/include/flamegpu/runtime/flamegpu_device_api.h b/include/flamegpu/runtime/flamegpu_device_api.h
index 638f974f1..abf75530c 100644
--- a/include/flamegpu/runtime/flamegpu_device_api.h
+++ b/include/flamegpu/runtime/flamegpu_device_api.h
@@ -65,6 +65,14 @@ class FLAMEGPU_READ_ONLY_DEVICE_API {
      */
     const DeviceEnvironment environment;
 
+    /**
+     * Access the current stepCount
+     * @return the current step count, 0 indexed unsigned.
+     */
+    __forceinline__ __device__ unsigned int getStepCounter() const {
+        return environment.get<unsigned int>("_stepCount");
+    }
+
  protected:
     Curve::NamespaceHash agent_func_name_hash;
 
diff --git a/include/flamegpu/runtime/flamegpu_host_api.h b/include/flamegpu/runtime/flamegpu_host_api.h
index 962fdad3f..da1f44775 100644
--- a/include/flamegpu/runtime/flamegpu_host_api.h
+++ b/include/flamegpu/runtime/flamegpu_host_api.h
@@ -75,6 +75,12 @@ class FLAMEGPU_HOST_API {
      */
     const HostEnvironment environment;
 
+    /**
+     * Access the current stepCount
+     * @return the current step count, 0 indexed unsigned.
+     */
+    unsigned int getStepCounter() const;
+
  private:
     /**
      * Used internally for tracking what CUB has already calculated temp memory for
diff --git a/include/flamegpu/runtime/utility/DeviceEnvironment.cuh b/include/flamegpu/runtime/utility/DeviceEnvironment.cuh
index 55e3ae95f..f40978679 100644
--- a/include/flamegpu/runtime/utility/DeviceEnvironment.cuh
+++ b/include/flamegpu/runtime/utility/DeviceEnvironment.cuh
@@ -93,7 +93,9 @@ __device__ __forceinline__ T DeviceEnvironment::get(const char(&name)[N]) const
     } else {
         curve_internal::d_curve_error = Curve::DEVICE_ERROR_UNKNOWN_VARIABLE;
         assert(false);
-        return *reinterpret_cast<T*>(reinterpret_cast<void*>(&flamegpu_internal::c_deviceEnvErrorPattern));
+        T rtn;
+        memcpy(&rtn, &flamegpu_internal::c_deviceEnvErrorPattern, sizeof(T));
+        return rtn;
     }
 }
 template<typename T, unsigned int N>
@@ -108,7 +110,9 @@ __device__ __forceinline__ T DeviceEnvironment::get(const char(&name)[N], const
     } else {
         curve_internal::d_curve_error = Curve::DEVICE_ERROR_UNKNOWN_VARIABLE;
         assert(false);
-        return *reinterpret_cast<T*>(reinterpret_cast<void*>(&flamegpu_internal::c_deviceEnvErrorPattern));
+        T rtn;
+        memcpy(&rtn, &flamegpu_internal::c_deviceEnvErrorPattern, sizeof(T));
+        return rtn;
     }
 }
 
diff --git a/include/flamegpu/runtime/utility/EnvironmentManager.cuh b/include/flamegpu/runtime/utility/EnvironmentManager.cuh
index 838c81f6c..505fa7043 100644
--- a/include/flamegpu/runtime/utility/EnvironmentManager.cuh
+++ b/include/flamegpu/runtime/utility/EnvironmentManager.cuh
@@ -22,6 +22,7 @@
 
 class EnvironmentDescription;
 class CUDAAgentModel;
+class CUDAAgent;
 
 
 /**
@@ -33,6 +34,10 @@ class CUDAAgentModel;
  * @note Not thread-safe
  */
 class EnvironmentManager {
+    /**
+     * Uses instance to for RTC compilation
+     */
+    friend class CUDAAgent;
     /**
      * Uses instance to initialise a models environment properties on the device
      */
@@ -148,8 +153,10 @@ class EnvironmentManager {
     /**
      * RTC functions hold thier own unique constants for environment variables. This function copies all environment variable to the RTC copies.
      * It can not be incorporated into init() as init will be called before RTC functions have been compiled.
+     * Uses the already populated Environment data from the cuda_model rather than environmentDescription.
+     * @param cuda_model the cuda model being initialised.
      */
-    void initRTC(const CUDAAgentModel& cuda_model, const EnvironmentDescription& desc);
+    void initRTC(const CUDAAgentModel& cuda_model);
     /**
      * Deactives all environmental properties linked to the named model from constant cache
      * @param model_name Name of the model
diff --git a/src/flamegpu/gpu/CUDAAgent.cu b/src/flamegpu/gpu/CUDAAgent.cu
index fb418df9a..c85980a71 100644
--- a/src/flamegpu/gpu/CUDAAgent.cu
+++ b/src/flamegpu/gpu/CUDAAgent.cu
@@ -219,8 +219,8 @@ void CUDAAgent::mapRuntimeVariables(const AgentFunctionData& func, const std::st
             agent_description.name.c_str(), func.initial_state.c_str());
     }
 
-    const Curve::VariableHash agent_hash = Curve::getInstance().variableRuntimeHash(agent_description.name.c_str());
-    const Curve::VariableHash func_hash = Curve::getInstance().variableRuntimeHash(func.name.c_str());
+    const Curve::VariableHash agent_hash = Curve::variableRuntimeHash(agent_description.name.c_str());
+    const Curve::VariableHash func_hash = Curve::variableRuntimeHash(func.name.c_str());
     const unsigned int agent_count = this->getStateSize(func.initial_state);
     // loop through the agents variables to map each variable name using cuRVE
     for (const auto &mmp : agent_description.variables) {
@@ -228,7 +228,7 @@ void CUDAAgent::mapRuntimeVariables(const AgentFunctionData& func, const std::st
         void* d_ptr = sm->second->getAgentListVariablePointer(mmp.first);
 
         // map using curve
-        const Curve::VariableHash var_hash = Curve::getInstance().variableRuntimeHash(mmp.first.c_str());
+        const Curve::VariableHash var_hash = Curve::variableRuntimeHash(mmp.first.c_str());
 
         // get the agent variable size
         const size_t type_size = mmp.second.type_size * mmp.second.elements;
@@ -271,15 +271,15 @@ void CUDAAgent::unmapRuntimeVariables(const AgentFunctionData& func) const {
             agent_description.name.c_str(), func.initial_state.c_str());
     }
 
-    const Curve::VariableHash agent_hash = Curve::getInstance().variableRuntimeHash(agent_description.name.c_str());
-    const Curve::VariableHash func_hash = Curve::getInstance().variableRuntimeHash(func.name.c_str());
+    const Curve::VariableHash agent_hash = Curve::variableRuntimeHash(agent_description.name.c_str());
+    const Curve::VariableHash func_hash = Curve::variableRuntimeHash(func.name.c_str());
     // loop through the agents variables to map each variable name using cuRVE
     for (const auto &mmp : agent_description.variables) {
         // get a device pointer for the agent variable name
         // void* d_ptr = sm->second->getAgentListVariablePointer(mmp.first);
 
         // unmap using curve
-        const Curve::VariableHash var_hash = Curve::getInstance().variableRuntimeHash(mmp.first.c_str());
+        const Curve::VariableHash var_hash = Curve::variableRuntimeHash(mmp.first.c_str());
         Curve::getInstance().unregisterVariableByHash(var_hash + agent_hash + func_hash);
     }
 
@@ -434,15 +434,15 @@ void CUDAAgent::mapNewRuntimeVariables(const AgentFunctionData& func, const unsi
             agent_description.name.c_str(), func.agent_output_state.c_str());
     }
 
-    const Curve::VariableHash _agent_birth_hash = Curve::getInstance().variableRuntimeHash("_agent_birth");
-    const Curve::VariableHash func_hash = Curve::getInstance().variableRuntimeHash(func.name.c_str());
+    const Curve::VariableHash _agent_birth_hash = Curve::variableRuntimeHash("_agent_birth");
+    const Curve::VariableHash func_hash = Curve::variableRuntimeHash(func.name.c_str());
     // loop through the agents variables to map each variable name using cuRVE
     for (const auto &mmp : agent_description.variables) {
         // get a device pointer for the agent variable name
         void* d_ptr = sm->second->getAgentNewListVariablePointer(mmp.first);
 
         // map using curve
-        const Curve::VariableHash var_hash = Curve::getInstance().variableRuntimeHash(mmp.first.c_str());
+        const Curve::VariableHash var_hash = Curve::variableRuntimeHash(mmp.first.c_str());
 
         // get the agent variable size
         size_t type_size = mmp.second.type_size * mmp.second.elements;
@@ -464,15 +464,15 @@ void CUDAAgent::mapNewRuntimeVariables(const AgentFunctionData& func, const unsi
 }
 
 void CUDAAgent::unmapNewRuntimeVariables(const AgentFunctionData& func) const {
-    const Curve::VariableHash _agent_birth_hash = Curve::getInstance().variableRuntimeHash("_agent_birth");
-    const Curve::VariableHash func_hash = Curve::getInstance().variableRuntimeHash(func.name.c_str());
+    const Curve::VariableHash _agent_birth_hash = Curve::variableRuntimeHash("_agent_birth");
+    const Curve::VariableHash func_hash = Curve::variableRuntimeHash(func.name.c_str());
     // loop through the agents variables to map each variable name using cuRVE
     for (const auto &mmp : agent_description.variables) {
         // get a device pointer for the agent variable name
         // void* d_ptr = sm->second->getAgentListVariablePointer(mmp.first);
 
         // unmap using curve
-        const Curve::VariableHash var_hash = Curve::getInstance().variableRuntimeHash(mmp.first.c_str());
+        const Curve::VariableHash var_hash = Curve::variableRuntimeHash(mmp.first.c_str());
         Curve::getInstance().unregisterVariableByHash(var_hash + _agent_birth_hash + func_hash);
 
         // no need to unmap RTC variables
@@ -575,8 +575,8 @@ void CUDAAgent::addInstantitateRTCFunction(const AgentFunctionData& func, bool f
     // curve rtc header
     CurveRTCHost curve_header;
     // agent function hash
-    Curve::NamespaceHash agentname_hash = Curve::getInstance().variableRuntimeHash(this->getAgentDescription().name.c_str());
-    Curve::NamespaceHash funcname_hash = Curve::getInstance().variableRuntimeHash(func.name.c_str());
+    Curve::NamespaceHash agentname_hash = Curve::variableRuntimeHash(this->getAgentDescription().name.c_str());
+    Curve::NamespaceHash funcname_hash = Curve::variableRuntimeHash(func.name.c_str());
     Curve::NamespaceHash agent_func_name_hash = agentname_hash + funcname_hash;
 
     // set agent function variables in rtc curve
@@ -589,7 +589,7 @@ void CUDAAgent::addInstantitateRTCFunction(const AgentFunctionData& func, bool f
         // Set input message variables in curve
         if (auto im = func.message_input.lock()) {
             // get the message input hash
-            Curve::NamespaceHash msg_in_hash = Curve::getInstance().variableRuntimeHash(im->name.c_str());
+            Curve::NamespaceHash msg_in_hash = Curve::variableRuntimeHash(im->name.c_str());
             for (auto msg_in_var : im->variables) {
                 // register message variables using combined hash
                 curve_header.registerVariable(msg_in_var.first.c_str(), msg_in_hash + agent_func_name_hash, msg_in_var.second.type.name(), msg_in_var.second.elements, true, false);
@@ -598,7 +598,7 @@ void CUDAAgent::addInstantitateRTCFunction(const AgentFunctionData& func, bool f
         // Set output message variables in curve
         if (auto om = func.message_output.lock()) {
             // get the message input hash
-            Curve::NamespaceHash msg_out_hash = Curve::getInstance().variableRuntimeHash(om->name.c_str());
+            Curve::NamespaceHash msg_out_hash = Curve::variableRuntimeHash(om->name.c_str());
             for (auto msg_out_var : om->variables) {
                 // register message variables using combined hash
                 curve_header.registerVariable(msg_out_var.first.c_str(), msg_out_hash + agent_func_name_hash, msg_out_var.second.type.name(), msg_out_var.second.elements, false, true);
@@ -607,17 +607,23 @@ void CUDAAgent::addInstantitateRTCFunction(const AgentFunctionData& func, bool f
         // Set agent output variables in curve
         if (auto ao = func.agent_output.lock()) {
             // get the message input hash
-            Curve::NamespaceHash agent_out_hash = Curve::getInstance().variableRuntimeHash("_agent_birth");
+            Curve::NamespaceHash agent_out_hash = Curve::variableRuntimeHash("_agent_birth");
             for (auto agent_out_var : ao->variables) {
                 // register message variables using combined hash
                 curve_header.registerVariable(agent_out_var.first.c_str(), agent_out_hash + funcname_hash, agent_out_var.second.type.name(), agent_out_var.second.elements, false, true);
             }
         }
     }
+
     // Set Environment variables in curve
-    Curve::NamespaceHash model_hash = Curve::getInstance().variableRuntimeHash(cuda_model.getModelDescription().name.c_str());
-    for (auto prop : cuda_model.getModelDescription().environment->getPropertiesMap()) {
-        curve_header.registerEnvVariable(prop.first.c_str(), model_hash, prop.second.type.name(), prop.second.elements);
+    Curve::NamespaceHash model_hash = Curve::variableRuntimeHash(cuda_model.getModelDescription().name.c_str());
+    for (auto p : EnvironmentManager::getInstance().getPropertiesMap()) {
+        if (p.first.first == cuda_model.getModelDescription().name) {
+            const char* variableName = p.first.second.c_str();
+            const char* type = p.second.type.name();
+            unsigned int elements = p.second.elements;
+            curve_header.registerEnvVariable(variableName, model_hash, type, elements);
+        }
     }
 
     // get the dynamically generated header from curve rtc
diff --git a/src/flamegpu/gpu/CUDAAgentModel.cu b/src/flamegpu/gpu/CUDAAgentModel.cu
index 1bb8efcd5..20c57294f 100644
--- a/src/flamegpu/gpu/CUDAAgentModel.cu
+++ b/src/flamegpu/gpu/CUDAAgentModel.cu
@@ -70,8 +70,6 @@ bool CUDAAgentModel::step() {
         fprintf(stdout, "Processing Simulation Step %u\n", step_count);
     }
 
-    step_count++;
-
     unsigned int nStreams = 1;
     std::string message_name;
     Curve::NamespaceHash message_name_inp_hash = 0;
@@ -482,10 +480,12 @@ bool CUDAAgentModel::step() {
 #ifdef VISUALISATION
             if (visualisation) {
                 NVTX_PUSH("CUDAAgentModel::step::ExitConditions::UpdateVisualisation");
-                visualisation->updateBuffers(step_count);
+                visualisation->updateBuffers(step_count+1);
                 NVTX_POP();
             }
 #endif
+            // If there were any exit conditions, we also need to update the step count
+            incrementStepCounter();
             return false;
         }
     // If we have exit conditions functions, we might have host agent creation
@@ -496,10 +496,12 @@ bool CUDAAgentModel::step() {
 #ifdef VISUALISATION
         if (visualisation) {
             NVTX_PUSH("CUDAAgentModel::step::UpdateVisualisation");
-            visualisation->updateBuffers(step_count);
+            visualisation->updateBuffers(step_count+1);
             NVTX_POP();
         }
 #endif
+    // Update step count at the end of the step - when it has completed.
+    incrementStepCounter();
     return true;
 }
 
@@ -774,6 +776,8 @@ void CUDAAgentModel::initialiseSingletons() {
 
         // Populate the environment properties in constant Cache
         singletons->environment.init(model->name, *model->environment);
+        // Add the CUDAAgentModel specific variables(s)
+        singletons->environment.add({model->name, "_stepCount"}, 0u, false);
 
         // Reinitialise random for this simulation instance
         singletons->rng.reseed(getSimulationConfig().random_seed);
@@ -781,7 +785,7 @@ void CUDAAgentModel::initialiseSingletons() {
         singletonsInitialised = true;
     }
 
-    // init RTC
+    // Ensure RTC is set up.
     initialiseRTC();
 }
 
@@ -809,7 +813,7 @@ void CUDAAgentModel::initialiseRTC() {
         }
 
         // Initialise device environment for RTC
-        singletons->environment.initRTC(*this, *model->environment);
+        singletons->environment.initRTC(*this);
 
         rtcInitialised = true;
     }
@@ -961,3 +965,7 @@ void CUDAAgentModel::RTCSetEnvironmentVariable(const char* variable_name, const
     }
 }
 
+void CUDAAgentModel::incrementStepCounter() {
+    this->step_count++;
+    this->singletons->environment.set({model->name, "_stepCount"}, this->step_count);
+}
diff --git a/src/flamegpu/runtime/flamegpu_host_api.cu b/src/flamegpu/runtime/flamegpu_host_api.cu
index 69fae3bcd..245322cfa 100644
--- a/src/flamegpu/runtime/flamegpu_host_api.cu
+++ b/src/flamegpu/runtime/flamegpu_host_api.cu
@@ -88,3 +88,13 @@ void FLAMEGPU_HOST_API::resizeTempStorage(const CUB_Config &cc, const unsigned i
     cub_largestAllocatedOp[cc] = items;
 }
 
+
+
+/**
+ * Access the current stepCount
+ * Sepearate implementation to avoid dependency loop with cuda agent model.
+ * @return the current step count, 0 indexed unsigned.
+ */
+unsigned int FLAMEGPU_HOST_API::getStepCounter() const {
+    return agentModel.getStepCounter();
+}
diff --git a/src/flamegpu/runtime/utility/EnvironmentManager.cu b/src/flamegpu/runtime/utility/EnvironmentManager.cu
index 92c969c45..6bb8bb83b 100644
--- a/src/flamegpu/runtime/utility/EnvironmentManager.cu
+++ b/src/flamegpu/runtime/utility/EnvironmentManager.cu
@@ -71,7 +71,7 @@ void EnvironmentManager::init(const std::string& model_name, const EnvironmentDe
     defragment(&orderedProperties);
 }
 
-void EnvironmentManager::initRTC(const CUDAAgentModel& cuda_model, const EnvironmentDescription& desc) {
+void EnvironmentManager::initRTC(const CUDAAgentModel& cuda_model) {
     // check to ensure that model name is not already registered
     auto res = cuda_agent_models.find(cuda_model.getModelDescription().name);
     if (res != cuda_agent_models.end()) {
@@ -80,10 +80,15 @@ void EnvironmentManager::initRTC(const CUDAAgentModel& cuda_model, const Environ
     // register model name
     cuda_agent_models.emplace(cuda_model.getModelDescription().name, cuda_model);
 
-    // loop through environment properties
-    for (auto p : desc.getPropertiesMap()) {
-        // Register variable for use in any RTC functions
-        cuda_model.RTCSetEnvironmentVariable(p.first.c_str(), p.second.data.ptr , p.second.data.length);
+    // loop through environment properties, already registered by cuda_
+    for (auto &p : properties) {
+        if (p.first.first == cuda_model.getModelDescription().name) {
+            auto var_name = p.first.second;
+            auto src = hc_buffer + p.second.offset;
+            auto length = p.second.length;
+            // Register variable for use in any RTC functions
+            cuda_model.RTCSetEnvironmentVariable(var_name.c_str(), src, length);
+        }
     }
 }
 
diff --git a/tests/CMakeLists.txt b/tests/CMakeLists.txt
index c95837d73..73f852824 100644
--- a/tests/CMakeLists.txt
+++ b/tests/CMakeLists.txt
@@ -106,6 +106,7 @@ SET(TEST_CASE_SRC
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_device_agent_creation.cu
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_device_api.cu
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_environment_manager.cu
+    ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_host_api.cu
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_host_agent_creation.cu
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_host_environment.cu
     ${CMAKE_CURRENT_SOURCE_DIR}/test_cases/runtime/test_host_random.cu
diff --git a/tests/test_cases/runtime/test_device_api.cu b/tests/test_cases/runtime/test_device_api.cu
index 195a2410a..838feb137 100644
--- a/tests/test_cases/runtime/test_device_api.cu
+++ b/tests/test_cases/runtime/test_device_api.cu
@@ -210,4 +210,89 @@ TEST(DeviceAPITest, ArrayUnsuitable) {
     }
 }
 
+// Test device_api::getStepCounter()
+FLAMEGPU_AGENT_FUNCTION(agent_testGetStepCounter, MsgNone, MsgNone) {
+    FLAMEGPU->setVariable<unsigned int>("step", FLAMEGPU->getStepCounter());
+    return ALIVE;
+}
+TEST(DeviceAPITest, getStepCounter) {
+    ModelDescription model("model");
+    AgentDescription &agent = model.newAgent("agent");
+    agent.newVariable<unsigned int>("step");
+    // Do nothing, but ensure variables are made available on device
+    AgentFunctionDescription &func = agent.newFunction("some_function", agent_testGetStepCounter);
+    model.newLayer().addAgentFunction(func);
+    // Init pop
+    const unsigned int agentCount = 1;
+    AgentPopulation init_population(agent, agentCount);
+    for (int i = 0; i< static_cast<int>(agentCount); i++) {
+        AgentInstance instance = init_population.getNextInstance("default");
+        instance.setVariable<unsigned int>("step", 0);
+    }
+    // Setup Model
+    CUDAAgentModel cuda_model(model);
+    cuda_model.setPopulationData(init_population);
+
+    const unsigned int STEPS = 2;
+    for (unsigned int step = 0; step < STEPS; step++) {
+        cuda_model.step();
+        // Recover data from device
+        AgentPopulation population(agent);
+        cuda_model.getPopulationData(population);
+        // Check data is correct.
+        for (unsigned int i = 0; i < population.getCurrentListSize(); i++) {
+            AgentInstance instance = population.getInstanceAt(i);
+            // Check neighbouring vars are correct
+            EXPECT_EQ(instance.getVariable<unsigned int>("step"), step);
+        }
+    }
+}
+
+// Only run the function based on the step accessed in the function condition.
+// I.e. only run the function once.
+FLAMEGPU_AGENT_FUNCTION_CONDITION(condition_testGetStepCounter) {
+    return FLAMEGPU->getStepCounter() == 0;
+}
+FLAMEGPU_AGENT_FUNCTION(condition_testGetStepCounterFunction, MsgNone, MsgNone) {
+    // Increment the counter of the number of times the agent ran the function.
+    unsigned int count = FLAMEGPU->getVariable<unsigned int>("count");
+    FLAMEGPU->setVariable<unsigned int>("count", count + 1);
+    return ALIVE;
+}
+TEST(DeviceAPITest, getStepCounterFunctionCondition) {
+    ModelDescription model("model");
+    AgentDescription &agent = model.newAgent("agent");
+    agent.newVariable<unsigned int>("count");
+    // Do nothing, but ensure variables are made available on device
+    AgentFunctionDescription &func = agent.newFunction("some_function", condition_testGetStepCounterFunction);
+    func.setFunctionCondition(condition_testGetStepCounter);
+    model.newLayer().addAgentFunction(func);
+    // Init pop
+    const unsigned int agentCount = 1;
+    AgentPopulation init_population(agent, agentCount);
+    for (int i = 0; i< static_cast<int>(agentCount); i++) {
+        AgentInstance instance = init_population.getNextInstance("default");
+        instance.setVariable<unsigned int>("count", 0);
+    }
+    const unsigned int STEPS = 4;
+    // Setup Model
+    CUDAAgentModel cuda_model(model);
+    cuda_model.setPopulationData(init_population);
+
+    // RUN STEPS steps of simulation.
+    cuda_model.SimulationConfig().steps = STEPS;
+    cuda_model.simulate();
+
+    // Recover data from device
+    AgentPopulation population(agent);
+    cuda_model.getPopulationData(population);
+    // Check data is correct.
+    const unsigned int EXPECTED_COUNT = 1;
+    for (unsigned int i = 0; i < population.getCurrentListSize(); i++) {
+        AgentInstance instance = population.getInstanceAt(i);
+        // Check neighbouring vars are correct
+        EXPECT_EQ(instance.getVariable<unsigned int>("count"), EXPECTED_COUNT);
+    }
+}
+
 }  // namespace test_device_api
diff --git a/tests/test_cases/runtime/test_environment_manager.cu b/tests/test_cases/runtime/test_environment_manager.cu
index 7ecbb6699..6aaf3c04b 100644
--- a/tests/test_cases/runtime/test_environment_manager.cu
+++ b/tests/test_cases/runtime/test_environment_manager.cu
@@ -114,13 +114,6 @@ TEST_F(EnvironmentManagerTest, OutOfMemory1) {
     ms->env.add<char, EnvironmentManager::MAX_BUFFER_SIZE / 2>("char_5kb_c", char_5kb_c);
     EXPECT_THROW(ms->run(), OutOfMemory);
 }
-TEST_F(EnvironmentManagerTest, OutOfMemory2) {
-    std::array<char, EnvironmentManager::MAX_BUFFER_SIZE / 2> char_5kb_a;
-    std::array<char, EnvironmentManager::MAX_BUFFER_SIZE / 2> char_5kb_b;
-    ms->env.add<char, EnvironmentManager::MAX_BUFFER_SIZE / 2>("char_5kb_a", char_5kb_a);
-    ms->env.add<char, EnvironmentManager::MAX_BUFFER_SIZE / 2>("char_5kb_b", char_5kb_b);
-    EXPECT_NO_THROW(ms->run());
-}
 
 // Multiple models
 TEST(EnvironmentManagerTest2, MultipleModels) {
diff --git a/tests/test_cases/runtime/test_host_api.cu b/tests/test_cases/runtime/test_host_api.cu
new file mode 100644
index 000000000..0bcea882b
--- /dev/null
+++ b/tests/test_cases/runtime/test_host_api.cu
@@ -0,0 +1,71 @@
+#include "gtest/gtest.h"
+
+#include "flamegpu/flame_api.h"
+#include "flamegpu/runtime/flamegpu_api.h"
+
+
+namespace test_host_api {
+
+// Test host_api::getStepCounter()
+
+FLAMEGPU_AGENT_FUNCTION(agent_function_getStepCounter, MsgNone, MsgNone) {
+    FLAMEGPU->setVariable<unsigned int>("step", FLAMEGPU->getStepCounter());
+    return ALIVE;
+}
+
+const unsigned int TOTAL_STEPS = 4;
+
+// globally scoped variable to track the value in the test?
+unsigned int expectedStepCounter = 0;
+
+// Init should always be 0th iteration/step
+FLAMEGPU_INIT_FUNCTION(init_testGetStepCounter) {
+    EXPECT_EQ(FLAMEGPU->getStepCounter(), 0);
+}
+// host is during, so 0? - @todo dynamic
+FLAMEGPU_HOST_FUNCTION(host_testGetStepCounter) {
+    EXPECT_EQ(FLAMEGPU->getStepCounter(), expectedStepCounter);
+}
+// Step functions are at the end of the step
+FLAMEGPU_STEP_FUNCTION(step_testGetStepCounter) {
+    EXPECT_EQ(FLAMEGPU->getStepCounter(), expectedStepCounter);
+}
+
+// Runs between steps - i.e. after step functions
+FLAMEGPU_EXIT_CONDITION(exitCondition_testGetStepCounter) {
+    EXPECT_EQ(FLAMEGPU->getStepCounter(), expectedStepCounter);
+    // Increment the counter used for testing multiple steps.
+    expectedStepCounter++;
+    return CONTINUE;
+}
+
+// exit is after all iterations, so stepCounter is the total number executed.
+FLAMEGPU_EXIT_FUNCTION(exit_testGetStepCounter) {
+    EXPECT_EQ(FLAMEGPU->getStepCounter(), TOTAL_STEPS);
+}
+
+TEST(hostAPITest, getStepCounter) {
+    ModelDescription model("model");
+    AgentDescription &agent = model.newAgent("agent");
+
+    model.addInitFunction(init_testGetStepCounter);
+    model.newLayer().addHostFunction(host_testGetStepCounter);
+    model.addStepFunction(step_testGetStepCounter);
+    model.addExitCondition(exitCondition_testGetStepCounter);
+    model.addExitFunction(exit_testGetStepCounter);
+
+    // Init pop
+    const unsigned int agentCount = 1;
+    AgentPopulation init_population(agent, agentCount);
+    for (int i = 0; i< static_cast<int>(agentCount); i++) {
+        AgentInstance instance = init_population.getNextInstance("default");
+    }
+    // Setup Model
+    CUDAAgentModel cuda_model(model);
+    cuda_model.setPopulationData(init_population);
+
+    cuda_model.SimulationConfig().steps = TOTAL_STEPS;
+    cuda_model.simulate();
+}
+
+}  // namespace test_host_api
diff --git a/tests/test_cases/runtime/test_rtc_device_api.cu b/tests/test_cases/runtime/test_rtc_device_api.cu
index f3c7af23e..63ff868c8 100644
--- a/tests/test_cases/runtime/test_rtc_device_api.cu
+++ b/tests/test_cases/runtime/test_rtc_device_api.cu
@@ -715,5 +715,47 @@ TEST(DeviceRTCAPITest, AgentFunction_cond_rtc) {
 }
 
 
+/**
+ * Test device_api::getStepCounter() in RTC code.
+ */
+const char* rtc_testGetStepCounter = R"###(
+FLAMEGPU_AGENT_FUNCTION(rtc_testGetStepCounter, MsgNone, MsgNone) {
+    FLAMEGPU->setVariable<unsigned int>("step", FLAMEGPU->getStepCounter());
+    return ALIVE;
+}
+)###";
+TEST(DeviceRTCAPITest, getStepCounter) {
+    ModelDescription model("model");
+    AgentDescription &agent = model.newAgent("agent");
+    agent.newVariable<unsigned int>("step");
+    // Do nothing, but ensure variables are made available on device
+    AgentFunctionDescription& func = agent.newRTCFunction("rtc_testGetStepCounter", rtc_testGetStepCounter);
+    model.newLayer().addAgentFunction(func);
+    // Init pop
+    const unsigned int agentCount = 1;
+    AgentPopulation init_population(agent, agentCount);
+    for (int i = 0; i< static_cast<int>(agentCount); i++) {
+        AgentInstance instance = init_population.getNextInstance("default");
+        instance.setVariable<unsigned int>("step", 0);
+    }
+    // Setup Model
+    CUDAAgentModel cuda_model(model);
+    cuda_model.setPopulationData(init_population);
+
+    const unsigned int STEPS = 2;
+    for (unsigned int step = 0; step < STEPS; step++) {
+        cuda_model.step();
+        // Recover data from device
+        AgentPopulation population(agent);
+        cuda_model.getPopulationData(population);
+        // Check data is correct.
+        for (unsigned int i = 0; i < population.getCurrentListSize(); i++) {
+            AgentInstance instance = population.getInstanceAt(i);
+            // Check neighbouring vars are correct
+            EXPECT_EQ(instance.getVariable<unsigned int>("step"), step);
+        }
+    }
+}
+
 
 }  // namespace test_rtc_device_api