Skip to content

Commit

Permalink
HostEnv import/export macro property.
Browse files Browse the repository at this point in the history
This special method also supports raw binary IO, for exceptionally large macro propertys unsuitable for xml/json.

Added an associated test in suite IOTest3 for each of the formats.
  • Loading branch information
Robadob authored and mondus committed Sep 22, 2023
1 parent da95c0f commit 8226cae
Show file tree
Hide file tree
Showing 12 changed files with 310 additions and 14 deletions.
2 changes: 1 addition & 1 deletion include/flamegpu/io/JSONStateWriter.h
Original file line number Diff line number Diff line change
Expand Up @@ -52,7 +52,7 @@ class JSONStateWriter : public StateWriter {
void writeConfig(const Simulation *sim_instance) override;
void writeStats(unsigned int iterations) override;
void writeEnvironment(const std::shared_ptr<const detail::EnvironmentManager>& env_manager) override;
void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env) override;
void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env, std::initializer_list<std::string> filter = {}) override;
void writeAgents(const util::StringPairUnorderedMap<std::shared_ptr<const AgentVector>>& agents_map) override;

private:
Expand Down
7 changes: 6 additions & 1 deletion include/flamegpu/io/StateWriter.h
Original file line number Diff line number Diff line change
Expand Up @@ -81,7 +81,12 @@ class StateWriter {
virtual void writeConfig(const Simulation *sim_instance) = 0;
virtual void writeStats(unsigned int iterations) = 0;
virtual void writeEnvironment(const std::shared_ptr<const detail::EnvironmentManager>& env_manager) = 0;
virtual void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env) = 0;
/**
* Write the macro environment block
* @param macro_env The macro environment to pull properties from
* @param filter If provided, only named properties will be written. Note, if filter contains missing properties it will fail
*/
virtual void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env, std::initializer_list<std::string> filter = {}) = 0;
virtual void writeAgents(const util::StringPairUnorderedMap<std::shared_ptr<const AgentVector>>& agents_map) = 0;
};
} // namespace io
Expand Down
2 changes: 1 addition & 1 deletion include/flamegpu/io/XMLStateWriter.h
Original file line number Diff line number Diff line change
Expand Up @@ -46,7 +46,7 @@ class XMLStateWriter : public StateWriter {
void writeConfig(const Simulation *sim_instance) override;
void writeStats(unsigned int iterations) override;
void writeEnvironment(const std::shared_ptr<const detail::EnvironmentManager>& env_manager) override;
void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env) override;
void writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env, std::initializer_list<std::string> filter = {}) override;
void writeAgents(const util::StringPairUnorderedMap<std::shared_ptr<const AgentVector>>& agents_map) override;

private:
Expand Down
27 changes: 26 additions & 1 deletion include/flamegpu/runtime/environment/HostEnvironment.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -36,7 +36,9 @@ class HostEnvironment {
/**
* Constructor, to be called by HostAPI
*/
explicit HostEnvironment(unsigned int instance_id, const std::shared_ptr<detail::EnvironmentManager> &env, const std::shared_ptr<detail::CUDAMacroEnvironment>& _macro_env);
explicit HostEnvironment(CUDASimulation &_simulation, cudaStream_t _stream,
std::shared_ptr<detail::EnvironmentManager> env,
std::shared_ptr<detail::CUDAMacroEnvironment> _macro_env);
/**
* Provides access to EnvironmentManager singleton
*/
Expand All @@ -50,6 +52,14 @@ class HostEnvironment {
* This is used to augment all variable names
*/
const unsigned int instance_id;
/**
* The relevant simulation, required for importing macro properties
*/
CUDASimulation& simulation;
/**
* CUDAStream for memcpys
*/
const cudaStream_t stream;

public:
/**
Expand Down Expand Up @@ -155,6 +165,21 @@ class HostEnvironment {
template<typename T>
HostMacroProperty_swig<T> getMacroProperty_swig(const std::string& name) const;
#endif
/**
* Import macro property data from file
* @param property_name Name of the macro property to import
* @param file_path Path to file containing macro property data (.json, .xml, .bin)
* @note This method supports raw binary files (.bin)
*/
void importMacroProperty(const std::string& property_name, const std::string& file_path) const;
/**
* Export macro property data to file
* @param property_name Name of the macro property to import
* @param file_path Path to file to export macro property data (.json, .xml. bin)
* @param pretty_print Print in readable or minified format (if available)
* @note This method supports raw binary files (.bin)
*/
void exportMacroProperty(const std::string& property_name, const std::string& file_path, bool pretty_print = true) const;
};

/**
Expand Down
18 changes: 15 additions & 3 deletions include/flamegpu/runtime/environment/HostMacroProperty.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -33,14 +33,26 @@ struct HostMacroProperty_MetaData {
}
/**
* Download data
* @note Only works first time
* @see force_download()
*/
void download() {
if (!h_base_ptr) {
force_download();
}
}

/**
* Downloads data from device to host (allocating host buffer if required)
* Sets has_changed flag to false, as host is current to device
*/
void force_download() {
if (!h_base_ptr) {
h_base_ptr = static_cast<char*>(malloc(elements * type_size));
gpuErrchk(cudaMemcpyAsync(h_base_ptr, d_base_ptr, elements * type_size, cudaMemcpyDeviceToHost, stream));
gpuErrchk(cudaStreamSynchronize(stream));
has_changed = false;
}
gpuErrchk(cudaMemcpyAsync(h_base_ptr, d_base_ptr, elements * type_size, cudaMemcpyDeviceToHost, stream));
gpuErrchk(cudaStreamSynchronize(stream));
has_changed = false;
}
/**
* Upload data
Expand Down
7 changes: 7 additions & 0 deletions include/flamegpu/simulation/detail/CUDAMacroEnvironment.h
Original file line number Diff line number Diff line change
Expand Up @@ -162,6 +162,13 @@ class CUDAMacroEnvironment {
* Used for IO
*/
const std::map<std::string, MacroEnvProp>& getPropertiesMap() const;
/**
* Return the metadata, if setup, for the named macro property
*
* @param property_name Name of the macro property to load
* @return nullptr if not currently cached
*/
std::shared_ptr<HostMacroProperty_MetaData> getHostPropertyMetadata(const std::string property_name);

private:
const CUDASimulation& cudaSimulation;
Expand Down
10 changes: 9 additions & 1 deletion src/flamegpu/io/JSONStateWriter.cu
Original file line number Diff line number Diff line change
Expand Up @@ -208,7 +208,7 @@ void JSONStateWriter::writeEnvironment(const std::shared_ptr<const detail::Envir

environment_written = true;
}
void JSONStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env) {
void JSONStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env, std::initializer_list<std::string> filter) {
if (!writer) {
THROW exception::UnknownInternalError("beginWrite() must be called before writeMacroEnvironment(), in JSONStateWriter::writeMacroEnvironment()");
} else if (macro_environment_written) {
Expand All @@ -220,6 +220,12 @@ void JSONStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::
writer->StartObject();
if (macro_env) {
const std::map<std::string, detail::CUDAMacroEnvironment::MacroEnvProp>& m_properties = macro_env->getPropertiesMap();
for (const auto &_filter : filter) {
if (m_properties.find(_filter) == m_properties.end()) {
THROW exception::InvalidEnvProperty("Macro property '%s' specified in filter does not exist, in JSONStateWriter::writeMacroEnvironment()", _filter.c_str());
}
}
std::set<std::string> filter_set = filter;
// Calculate largest buffer in map
size_t max_len = 0;
for (const auto& [_, prop] : m_properties) {
Expand All @@ -230,6 +236,8 @@ void JSONStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::
char* const t_buffer = static_cast<char*>(malloc(max_len));
// Write out each array (all are written out as 1D arrays for simplicity given variable dimensions)
for (const auto& [name, prop] : m_properties) {
if (!filter_set.empty() && filter_set.find(name) == filter_set.end())
continue;
// Copy data
const size_t element_ct = std::accumulate(prop.elements.begin(), prop.elements.end(), 1, std::multiplies<unsigned int>());
gpuErrchk(cudaMemcpy(t_buffer, prop.d_ptr, element_ct * prop.type_size, cudaMemcpyDeviceToHost));
Expand Down
10 changes: 9 additions & 1 deletion src/flamegpu/io/XMLStateWriter.cu
Original file line number Diff line number Diff line change
Expand Up @@ -250,7 +250,7 @@ void XMLStateWriter::writeEnvironment(const std::shared_ptr<const detail::Enviro

environment_written = true;
}
void XMLStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env) {
void XMLStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::CUDAMacroEnvironment>& macro_env, std::initializer_list<std::string> filter) {
if (!doc || !pRoot) {
THROW exception::UnknownInternalError("beginWrite() must be called before writeMacroEnvironment(), in XMLStateWriter::writeMacroEnvironment()");
} else if (macro_environment_written) {
Expand All @@ -260,6 +260,12 @@ void XMLStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::C
tinyxml2::XMLElement *pElement = doc->NewElement("macro_environment");
if (macro_env) {
const std::map<std::string, detail::CUDAMacroEnvironment::MacroEnvProp>& m_properties = macro_env->getPropertiesMap();
for (const auto &_filter : filter) {
if (m_properties.find(_filter) == m_properties.end()) {
THROW exception::InvalidEnvProperty("Macro property '%s' specified in filter does not exist, in XMLStateWriter::writeMacroEnvironment()", _filter.c_str());
}
}
std::set<std::string> filter_set = filter;
// Calculate largest buffer in map
size_t max_len = 0;
for (const auto& [_, prop] : m_properties) {
Expand All @@ -270,6 +276,8 @@ void XMLStateWriter::writeMacroEnvironment(const std::shared_ptr<const detail::C
char* const t_buffer = static_cast<char*>(malloc(max_len));
// Write out each array (all are written out as 1D arrays for simplicity given variable dimensions)
for (const auto& [name, prop] : m_properties) {
if (!filter_set.empty() && filter_set.find(name) == filter_set.end())
continue;
// Copy data
const size_t element_ct = std::accumulate(prop.elements.begin(), prop.elements.end(), 1, std::multiplies<unsigned int>());
gpuErrchk(cudaMemcpy(t_buffer, prop.d_ptr, element_ct * prop.type_size, cudaMemcpyDeviceToHost));
Expand Down
2 changes: 1 addition & 1 deletion src/flamegpu/runtime/HostAPI.cu
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@ HostAPI::HostAPI(CUDASimulation &_agentModel,
const unsigned int _streamId,
cudaStream_t _stream)
: random(rng)
, environment(_agentModel.getInstanceID(), env, macro_env)
, environment(_agentModel, _stream, env, macro_env)
, agentModel(_agentModel)
, d_output_space(nullptr)
, d_output_space_size(0)
Expand Down
111 changes: 107 additions & 4 deletions src/flamegpu/runtime/environment/HostEnvironment.cu
Original file line number Diff line number Diff line change
@@ -1,10 +1,113 @@
#include "flamegpu/runtime/environment/HostEnvironment.cuh"

#include <filesystem>
#include <fstream>
#include <iterator>
#include <numeric>
#include <vector>

#include "flamegpu/io/StateWriter.h"
#include "flamegpu/io/StateWriterFactory.h"
#include "flamegpu/io/StateReader.h"
#include "flamegpu/io/StateReaderFactory.h"
#include "flamegpu/simulation/CUDASimulation.h"

namespace flamegpu {

HostEnvironment::HostEnvironment(const unsigned int _instance_id, const std::shared_ptr<detail::EnvironmentManager> &env, const std::shared_ptr<detail::CUDAMacroEnvironment>& _macro_env)
: env_mgr(env)
, macro_env(_macro_env)
, instance_id(_instance_id) { }
HostEnvironment::HostEnvironment(CUDASimulation &_simulation, cudaStream_t _stream,
std::shared_ptr<detail::EnvironmentManager> env,
std::shared_ptr<detail::CUDAMacroEnvironment> _macro_env)
: env_mgr(std::move(env))
, macro_env(std::move(_macro_env))
, instance_id(_simulation.getInstanceID())
, simulation(_simulation)
, stream(_stream) { }

void HostEnvironment::importMacroProperty(const std::string& property_name, const std::string& file_path) const {
// Validate the property exists
const auto &m_props = macro_env->getPropertiesMap();
const auto &m_prop = m_props.find(property_name);
if (m_prop == m_props.end()) {
THROW exception::InvalidEnvProperty("The environment macro property '%s' was not found within the model description, in HostEnvironment::importMacroProperty().", property_name.c_str());
}
const unsigned int m_prop_elements = std::accumulate(m_prop->second.elements.begin(), m_prop->second.elements.end(), 1, std::multiplies<unsigned int>());
try {
io::StateReader *read__ = io::StateReaderFactory::createReader(file_path);
read__->parse(file_path, simulation.getModelDescription().shared_from_this(), Verbosity::Quiet);
std::unordered_map<std::string, std::vector<char>> macro_init;
read__->getMacroEnvironment(macro_init);
// Validate the property exists within macro_init
const auto &l_prop = macro_init.find(property_name);
if (l_prop == macro_init.end()) {
THROW exception::InvalidEnvProperty("The environment macro property '%s' was not found within the input file '%s'.", property_name.c_str(), file_path.c_str());
}
// Check the length validates
if (l_prop->second.size() != m_prop_elements * m_prop->second.type_size) {
THROW exception::InvalidInputFile("Length of input file '%s's environment macro property '%s' does not match, (%u != %u), in HostEnvironment::importMacroProperty()",
file_path.c_str(), property_name.c_str(), static_cast<unsigned int>(l_prop->second.size()), static_cast<unsigned int>(m_prop_elements * m_prop->second.type_size));
}
gpuErrchk(cudaMemcpyAsync(m_prop->second.d_ptr, l_prop->second.data(), l_prop->second.size(), cudaMemcpyHostToDevice, stream));
} catch (const exception::UnsupportedFileType&) {
const std::string extension = std::filesystem::path(file_path).extension().string();
if (extension == ".bin") {
// Additionally support raw binary dump
// Read the file
std::ifstream input(file_path, std::ios::binary);
std::vector buffer(std::istreambuf_iterator<char>(input), {});
// Check the length validates
if (buffer.size() != m_prop_elements * m_prop->second.type_size) {
THROW exception::InvalidInputFile("Length of binary input file '%s' does not match the environment macro property '%s', (%u != %u), in HostEnvironment::importMacroProperty()",
file_path.c_str(), property_name.c_str(), static_cast<unsigned int>(buffer.size()), static_cast<unsigned int>(m_prop_elements * m_prop->second.type_size));
}
// Update the property
gpuErrchk(cudaMemcpyAsync(m_prop->second.d_ptr, buffer.data(), buffer.size(), cudaMemcpyHostToDevice, stream));
} else {
throw;
}
}
gpuErrchk(cudaStreamSynchronize(stream));
// If macro property exists in cache sync cache
if (const auto cache = macro_env->getHostPropertyMetadata(property_name)) {
cache->force_download();
}
}
void HostEnvironment::exportMacroProperty(const std::string& property_name, const std::string& file_path, bool pretty_print) const {
// If macro property exists in cache sync cache
if (const auto cache = macro_env->getHostPropertyMetadata(property_name)) {
cache->upload();
}
try {
io::StateWriter* write__ = io::StateWriterFactory::createWriter(file_path);
write__->beginWrite(file_path, pretty_print);
write__->writeMacroEnvironment(macro_env, { property_name });
write__->endWrite();
} catch (const exception::UnsupportedFileType&) {
const std::string extension = std::filesystem::path(file_path).extension().string();
if (extension == ".bin") {
// Additionally support raw binary dump
// Validate the property exists
const auto &m_props = macro_env->getPropertiesMap();
const auto &m_prop = m_props.find(property_name);
if (m_prop == m_props.end()) {
THROW exception::InvalidEnvProperty("The environment macro property '%s' was not found within the model description, in HostEnvironment::exportMacroProperty().", property_name.c_str());
}
// Check the file doesn't already exist
if (std::filesystem::exists(file_path)) {
THROW exception::FileAlreadyExists("The binary output file '%s' already exists, in HostEnvironment::exportMacroProperty().", file_path.c_str());
}
// Copy the data to a temporary buffer on host
const unsigned int m_prop_elements = std::accumulate(m_prop->second.elements.begin(), m_prop->second.elements.end(), 1, std::multiplies<unsigned int>());
std::vector<char> buffer;
buffer.resize(m_prop_elements * m_prop->second.type_size);
gpuErrchk(cudaMemcpyAsync(buffer.data(), m_prop->second.d_ptr, m_prop_elements * m_prop->second.type_size, cudaMemcpyDeviceToHost, stream));
gpuErrchk(cudaStreamSynchronize(stream));
// Output to file
std::ofstream output(file_path, std::ios::binary);
output.write(buffer.data(), buffer.size());
} else {
throw;
}
}
}

} // namespace flamegpu
8 changes: 8 additions & 0 deletions src/flamegpu/simulation/detail/CUDAMacroEnvironment.cu
Original file line number Diff line number Diff line change
Expand Up @@ -102,6 +102,14 @@ void CUDAMacroEnvironment::unmapRTCVariables(detail::curve::CurveRTCHost& curve_
const std::map<std::string, CUDAMacroEnvironment::MacroEnvProp>& CUDAMacroEnvironment::getPropertiesMap() const {
return properties;
}

std::shared_ptr<HostMacroProperty_MetaData> CUDAMacroEnvironment::getHostPropertyMetadata(const std::string property_name) {
auto cache = host_cache.find(property_name);
if (cache != host_cache.end()) {
return cache->second.lock();
}
return nullptr;
}
#if !defined(FLAMEGPU_SEATBELTS) || FLAMEGPU_SEATBELTS
void CUDAMacroEnvironment::resetFlagsAsync(const std::vector<cudaStream_t> &streams) {
unsigned int i = 0;
Expand Down
Loading

0 comments on commit 8226cae

Please sign in to comment.