arm_compute v18.05
diff --git a/tests/framework/Exceptions.h b/tests/framework/Exceptions.h
index f35c350..687305b 100644
--- a/tests/framework/Exceptions.h
+++ b/tests/framework/Exceptions.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -94,6 +94,10 @@
*/
LogLevel level() const;
+ /** Get the error message.
+ *
+ * @return error message.
+ */
const char *what() const noexcept override;
private:
diff --git a/tests/framework/Framework.cpp b/tests/framework/Framework.cpp
index 3091b66..fd0afe9 100644
--- a/tests/framework/Framework.cpp
+++ b/tests/framework/Framework.cpp
@@ -24,6 +24,9 @@
#include "Framework.h"
#include "support/ToolchainSupport.h"
+#ifdef ARM_COMPUTE_CL
+#include "arm_compute/runtime/CL/CLScheduler.h"
+#endif /* ARM_COMPUTE_CL */
#include <chrono>
#include <iostream>
@@ -59,6 +62,11 @@
_available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_US), Instrument::make_instrument<OpenCLTimer, ScaleFactor::TIME_US>);
_available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_MS), Instrument::make_instrument<OpenCLTimer, ScaleFactor::TIME_MS>);
_available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_S), Instrument::make_instrument<OpenCLTimer, ScaleFactor::TIME_S>);
+ _available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::NONE), Instrument::make_instrument<OpenCLMemoryUsage, ScaleFactor::NONE>);
+ _available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::SCALE_1K),
+ Instrument::make_instrument<OpenCLMemoryUsage, ScaleFactor::SCALE_1K>);
+ _available_instruments.emplace(std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::SCALE_1M),
+ Instrument::make_instrument<OpenCLMemoryUsage, ScaleFactor::SCALE_1M>);
#endif /* ARM_COMPUTE_CL */
}
@@ -288,6 +296,8 @@
try
{
+ profiler.test_start();
+
test_case->do_setup();
for(int i = 0; i < _num_iterations; ++i)
@@ -312,6 +322,8 @@
test_case->do_teardown();
+ profiler.test_stop();
+
// Change status to success if no error has happend
if(result.status == TestResult::Status::NOT_RUN)
{
@@ -508,7 +520,8 @@
const std::chrono::time_point<std::chrono::high_resolution_clock> start = std::chrono::high_resolution_clock::now();
- int id = 0;
+ int id = 0;
+ int id_run_test = 0;
for(auto &test_factory : _test_factories)
{
@@ -517,7 +530,21 @@
if(_test_filter.is_selected(test_info))
{
+#ifdef ARM_COMPUTE_CL
+ // Every 5000 tests, reset the OpenCL context to release the allocated memory
+ if((id_run_test % 5000) == 0)
+ {
+ cl::Context::setDefault(cl::Context());
+ CLScheduler::get().set_context(cl::Context());
+ CLKernelLibrary::get().clear_programs_cache();
+
+ cl::Context::setDefault(cl::Context(CL_DEVICE_TYPE_DEFAULT));
+ CLScheduler::get().set_context(cl::Context::getDefault());
+ }
+#endif // ARM_COMPUTE_CL
run_test(test_info, *test_factory);
+
+ ++id_run_test;
}
++id;
diff --git a/tests/framework/Framework.h b/tests/framework/Framework.h
index d7a9cfb..65ffc0a 100644
--- a/tests/framework/Framework.h
+++ b/tests/framework/Framework.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -65,10 +65,10 @@
*/
struct TestInfo
{
- int id;
- std::string name;
- DatasetMode mode;
- TestCaseFactory::Status status;
+ int id; /**< Test ID */
+ std::string name; /**< Test name */
+ DatasetMode mode; /**< Test data set mode */
+ TestCaseFactory::Status status; /**< Test status */
};
inline bool operator<(const TestInfo &lhs, const TestInfo &rhs)
diff --git a/tests/framework/Profiler.cpp b/tests/framework/Profiler.cpp
index 646c665..7b95279 100644
--- a/tests/framework/Profiler.cpp
+++ b/tests/framework/Profiler.cpp
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -37,6 +37,14 @@
_instruments.emplace_back(std::move(instrument));
}
+void Profiler::test_start()
+{
+ for(auto &instrument : _instruments)
+ {
+ instrument->test_start();
+ }
+}
+
void Profiler::start()
{
for(auto &instrument : _instruments)
@@ -47,14 +55,29 @@
void Profiler::stop()
{
- for(auto &instrument : _instruments)
+ for(auto instrument = _instruments.rbegin(); instrument != _instruments.rend(); instrument++)
{
- instrument->stop();
+ (*instrument)->stop();
+ }
+ for(const auto &instrument : _instruments)
+ {
+ for(const auto &measurement : instrument->measurements())
+ {
+ _measurements[instrument->id() + "/" + measurement.first].push_back(measurement.second);
+ }
+ }
+}
+
+void Profiler::test_stop()
+{
+ for(auto instrument = _instruments.rbegin(); instrument != _instruments.rend(); instrument++)
+ {
+ (*instrument)->test_stop();
}
for(const auto &instrument : _instruments)
{
- for(const auto &measurement : instrument->measurements())
+ for(const auto &measurement : instrument->test_measurements())
{
_measurements[instrument->id() + "/" + measurement.first].push_back(measurement.second);
}
diff --git a/tests/framework/Profiler.h b/tests/framework/Profiler.h
index 930075e..34c5224 100644
--- a/tests/framework/Profiler.h
+++ b/tests/framework/Profiler.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -57,13 +57,34 @@
*/
void add(std::unique_ptr<Instrument> instrument);
- /** Start all added instruments to measure performance. */
+ /** Call test_start() on all the added instruments
+ *
+ * Called before the test set up starts
+ */
+ void test_start();
+
+ /** Call start() on all the added instruments
+ *
+ * Called just before the run of the test starts
+ */
void start();
- /** Stop all added instruments. */
+ /** Call stop() on all the added instruments
+ *
+ * Called just after the run of the test ends
+ */
void stop();
- /** Return measurements for all instruments. */
+ /** Call test_stop() on all the added instruments
+ *
+ * Called after the test teardown ended
+ */
+ void test_stop();
+
+ /** Return measurements for all instruments.
+ *
+ * @return measurements for all instruments.
+ */
const MeasurementsMap &measurements() const;
private:
diff --git a/tests/framework/SConscript b/tests/framework/SConscript
index e740828..b8574bd 100644
--- a/tests/framework/SConscript
+++ b/tests/framework/SConscript
@@ -69,7 +69,7 @@
if not env['opencl']:
# Remove OpenCLTimer files
- files = [f for f in files if "OpenCLTimer" not in os.path.basename(str(f))]
+ files = [f for f in files if "OpenCL" not in os.path.basename(str(f))]
if not framework_env['mali']:
# Remove MALI files
diff --git a/tests/framework/TestCase.h b/tests/framework/TestCase.h
index 18dd12e..d7bf54d 100644
--- a/tests/framework/TestCase.h
+++ b/tests/framework/TestCase.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -40,9 +40,13 @@
class TestCase
{
public:
+ /** Setup the test */
virtual void do_setup() {};
+ /** Run the test */
virtual void do_run() {};
+ /** Sync the test */
virtual void do_sync() {};
+ /** Teardown the test */
virtual void do_teardown() {};
/** Default destructor. */
@@ -54,10 +58,15 @@
friend class TestCaseFactory;
};
+/** Data test case class */
template <typename T>
class DataTestCase : public TestCase
{
protected:
+ /** Construct a data test case.
+ *
+ * @param[in] data Test data.
+ */
explicit DataTestCase(T data)
: _data{ std::move(data) }
{
diff --git a/tests/framework/TestCaseFactory.h b/tests/framework/TestCaseFactory.h
index b8c8cdb..7164f8f 100644
--- a/tests/framework/TestCaseFactory.h
+++ b/tests/framework/TestCaseFactory.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -110,6 +110,7 @@
std::unique_ptr<TestCase> make() const override;
};
+/** Implementation of a test case factory to create data test cases. */
template <typename T, typename D>
class DataTestCaseFactory final : public TestCaseFactory
{
diff --git a/tests/framework/TestResult.h b/tests/framework/TestResult.h
index e71ef95..cdace17 100644
--- a/tests/framework/TestResult.h
+++ b/tests/framework/TestResult.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -71,8 +71,8 @@
{
}
- Status status{ Status::NOT_RUN }; //< Execution status
- Profiler::MeasurementsMap measurements{}; //< Profiling information
+ Status status{ Status::NOT_RUN }; /**< Execution status */
+ Profiler::MeasurementsMap measurements{}; /**< Profiling information */
};
} // namespace framework
} // namespace test
diff --git a/tests/framework/command_line/CommonOptions.h b/tests/framework/command_line/CommonOptions.h
index 2da2c99..651316c 100644
--- a/tests/framework/command_line/CommonOptions.h
+++ b/tests/framework/command_line/CommonOptions.h
@@ -40,15 +40,15 @@
enum class LogLevel;
/** Common command line options used to configure the framework
- *
- * The options in this object get populated when "parse()" is called on the parser used to construct it.
- * The expected workflow is:
- *
- * CommandLineParser parser;
- * CommonOptions options( parser );
- * parser.parse(argc, argv);
- * if(options.log_level->value() > LogLevel::NONE) --> Use the options values
- */
+ *
+ * The options in this object get populated when "parse()" is called on the parser used to construct it.
+ * The expected workflow is:
+ *
+ * CommandLineParser parser;
+ * CommonOptions options( parser );
+ * parser.parse(argc, argv);
+ * if(options.log_level->value() > LogLevel::NONE) --> Use the options values
+ */
class CommonOptions
{
public:
@@ -57,7 +57,9 @@
* @param[in,out] parser A parser on which "parse()" hasn't been called yet.
*/
CommonOptions(CommandLineParser &parser);
+ /** Prevent instances of this class from being copy constructed */
CommonOptions(const CommonOptions &) = delete;
+ /** Prevent instances of this class from being copied */
CommonOptions &operator=(const CommonOptions &) = delete;
/** Create the printers based on parsed command line options
*
@@ -67,19 +69,19 @@
*/
std::vector<std::unique_ptr<Printer>> create_printers();
- ToggleOption *help;
- EnumListOption<InstrumentsDescription> *instruments;
- SimpleOption<int> *iterations;
- SimpleOption<int> *threads;
- EnumOption<LogFormat> *log_format;
- SimpleOption<std::string> *log_file;
- EnumOption<LogLevel> *log_level;
- ToggleOption *throw_errors;
- ToggleOption *color_output;
- ToggleOption *pretty_console;
- SimpleOption<std::string> *json_file;
- SimpleOption<std::string> *pretty_file;
- std::vector<std::shared_ptr<std::ofstream>> log_streams;
+ ToggleOption *help; /**< Show help option */
+ EnumListOption<InstrumentsDescription> *instruments; /**< Instruments option */
+ SimpleOption<int> *iterations; /**< Number of iterations option */
+ SimpleOption<int> *threads; /**< Number of threads option */
+ EnumOption<LogFormat> *log_format; /**< Log format option */
+ SimpleOption<std::string> *log_file; /**< Log file option */
+ EnumOption<LogLevel> *log_level; /**< Logging level option */
+ ToggleOption *throw_errors; /**< Throw errors option */
+ ToggleOption *color_output; /**< Color output option */
+ ToggleOption *pretty_console; /**< Pretty console option */
+ SimpleOption<std::string> *json_file; /**< JSON output file option */
+ SimpleOption<std::string> *pretty_file; /**< Pretty output file option */
+ std::vector<std::shared_ptr<std::ofstream>> log_streams; /**< Log streams */
};
} // namespace framework
diff --git a/tests/framework/command_line/EnumListOption.h b/tests/framework/command_line/EnumListOption.h
index 6155a5d..39006d8 100644
--- a/tests/framework/command_line/EnumListOption.h
+++ b/tests/framework/command_line/EnumListOption.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -60,7 +60,12 @@
EnumListOption(std::string name, std::set<T> allowed_values, std::initializer_list<T> &&default_values);
bool parse(std::string value) override;
- std::string help() const override;
+ std::string help() const override;
+
+ /** Get the values of the option.
+ *
+ * @return a list of the selected option values.
+ */
const std::vector<T> &value() const;
private:
diff --git a/tests/framework/command_line/EnumOption.h b/tests/framework/command_line/EnumOption.h
index 1abba77..14d6185 100644
--- a/tests/framework/command_line/EnumOption.h
+++ b/tests/framework/command_line/EnumOption.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -59,7 +59,12 @@
bool parse(std::string value) override;
std::string help() const override;
- const T &value() const;
+
+ /** Get the selected value.
+ *
+ * @return get the selected enum value.
+ */
+ const T &value() const;
private:
std::set<T> _allowed_values{};
diff --git a/tests/framework/command_line/ListOption.h b/tests/framework/command_line/ListOption.h
index 8b1bb3d..07184e8 100644
--- a/tests/framework/command_line/ListOption.h
+++ b/tests/framework/command_line/ListOption.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -53,7 +53,12 @@
ListOption(std::string name, std::initializer_list<T> &&default_values);
bool parse(std::string value) override;
- std::string help() const override;
+ std::string help() const override;
+
+ /** Get the list of option values.
+ *
+ * @return get the list of option values.
+ */
const std::vector<T> &value() const;
private:
diff --git a/tests/framework/command_line/SimpleOption.h b/tests/framework/command_line/SimpleOption.h
index e6e8045..d02778e 100644
--- a/tests/framework/command_line/SimpleOption.h
+++ b/tests/framework/command_line/SimpleOption.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -50,9 +50,25 @@
*/
SimpleOption(std::string name, T default_value);
+ /** Parses the given string.
+ *
+ * @param[in] value String representation as passed on the command line.
+ *
+ * @return True if the value could be parsed by the specific subclass.
+ */
bool parse(std::string value) override;
+
+ /** Help message for the option.
+ *
+ * @return String representing the help message for the specific subclass.
+ */
std::string help() const override;
- const T &value() const;
+
+ /** Get the option value.
+ *
+ * @return the option value.
+ */
+ const T &value() const;
protected:
T _value{};
diff --git a/tests/framework/datasets/CartesianProductDataset.h b/tests/framework/datasets/CartesianProductDataset.h
index 438a782..b2790d7 100644
--- a/tests/framework/datasets/CartesianProductDataset.h
+++ b/tests/framework/datasets/CartesianProductDataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -64,6 +64,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
CartesianProductDataset(CartesianProductDataset &&) = default;
/** Type of the dataset. */
@@ -72,6 +73,11 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator.
+ *
+ * @param[in] dataset1 Dataset 1.
+ * @param[in] dataset2 Dataset 2.
+ */
iterator(const T_noref *dataset1, const U_noref *dataset2)
: _iter1{ dataset1->begin() },
_dataset2{ dataset2 },
@@ -79,23 +85,40 @@
{
}
+ /** Allow instances of this class to be copy constructed */
iterator(const iterator &) = default;
+ /** Allow instances of this class to be copied */
iterator &operator=(const iterator &) = default;
- iterator(iterator &&) = default;
+ /** Allow instances of this class to be move constructed */
+ iterator(iterator &&) = default;
+ /** Allow instances of this class to be moved */
iterator &operator=(iterator &&) = default;
+ /** Default destructor */
~iterator() = default;
+ /** Get the description of the current value.
+ *
+ * @return description of the current value.
+ */
std::string description() const
{
return _iter1.description() + ":" + _iter2.description();
}
+ /** Get the value of the iterator.
+ *
+ * @return the value of the iterator.
+ */
CartesianProductDataset::type operator*() const
{
return std::tuple_cat(*_iter1, *_iter2);
}
+ /** Inrement the iterator.
+ *
+ * @return *this;
+ */
iterator &operator++()
{
++_second_pos;
@@ -159,6 +182,13 @@
return CartesianProductDataset<T, U>(std::forward<T>(dataset1), std::forward<U>(dataset2));
}
+/** Helper function to create a @ref CartesianProductDataset.
+ *
+ * @param[in] dataset1 First dataset.
+ * @param[in] dataset2 Second dataset.
+ *
+ * @return A grid dataset.
+ */
template <typename T, typename U>
CartesianProductDataset<T, U>
operator*(T &&dataset1, U &&dataset2)
diff --git a/tests/framework/datasets/ContainerDataset.h b/tests/framework/datasets/ContainerDataset.h
index 80616c4..8dfd216 100644
--- a/tests/framework/datasets/ContainerDataset.h
+++ b/tests/framework/datasets/ContainerDataset.h
@@ -72,8 +72,10 @@
{
}
+ /** Allow instances of this class to be copy constructed */
ContainerDataset(const ContainerDataset &) = default;
- ContainerDataset(ContainerDataset &&) = default;
+ /** Allow instances of this class to be move constructed */
+ ContainerDataset(ContainerDataset &&) = default;
/** Type of the dataset. */
using type = std::tuple<container_value_type>;
@@ -81,22 +83,39 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct iterator
+ *
+ * @param[in] name Description of the values.
+ * @param[in] iterator Iterator for the values.
+ */
iterator(std::string name, container_const_iterator iterator)
: _name{ name }, _iterator{ iterator }
{
}
+ /** Get a description of the current value.
+ *
+ * @return a description.
+ */
std::string description() const
{
using support::cpp11::to_string;
return _name + "=" + to_string(*_iterator);
}
+ /** Get the current value.
+ *
+ * @return the current value.
+ */
ContainerDataset::type operator*() const
{
return std::make_tuple(*_iterator);
}
+ /** Increment the iterator.
+ *
+ * @return this.
+ */
iterator &operator++()
{
++_iterator;
diff --git a/tests/framework/datasets/Dataset.h b/tests/framework/datasets/Dataset.h
index d916730..5fcdc49 100644
--- a/tests/framework/datasets/Dataset.h
+++ b/tests/framework/datasets/Dataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -39,10 +39,13 @@
class Dataset
{
protected:
- Dataset() = default;
+ /** Default constructor. */
+ Dataset() = default;
+ /** Default destructor. */
~Dataset() = default;
public:
+ /** Allow instances of this class to be move constructed */
Dataset(Dataset &&) = default;
};
@@ -62,9 +65,11 @@
{
}
+ /** Default destructor. */
~NamedDataset() = default;
public:
+ /** Allow instances of this class to be move constructed */
NamedDataset(NamedDataset &&) = default;
/** Return name of the dataset.
diff --git a/tests/framework/datasets/InitializerListDataset.h b/tests/framework/datasets/InitializerListDataset.h
index 7d32234..f90e0b7 100644
--- a/tests/framework/datasets/InitializerListDataset.h
+++ b/tests/framework/datasets/InitializerListDataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -59,6 +59,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
InitializerListDataset(InitializerListDataset &&) = default;
/** Type of the dataset. */
@@ -67,22 +68,39 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator for the dataset
+ *
+ * @param[in] name Name of the dataset.
+ * @param[in] iterator Iterator of the dataset values.
+ */
iterator(std::string name, data_const_iterator iterator)
: _name{ name }, _iterator{ iterator }
{
}
+ /** Get a description of the current value.
+ *
+ * @return a description.
+ */
std::string description() const
{
using support::cpp11::to_string;
return _name + "=" + to_string(*_iterator);
}
+ /** Get the current value.
+ *
+ * @return the current value.
+ */
InitializerListDataset::type operator*() const
{
return std::make_tuple(*_iterator);
}
+ /** Increment the iterator.
+ *
+ * @return *this.
+ */
iterator &operator++()
{
++_iterator;
diff --git a/tests/framework/datasets/JoinDataset.h b/tests/framework/datasets/JoinDataset.h
index d682c19..bf504ec 100644
--- a/tests/framework/datasets/JoinDataset.h
+++ b/tests/framework/datasets/JoinDataset.h
@@ -64,6 +64,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
JoinDataset(JoinDataset &&) = default;
/** Type of the dataset. */
@@ -72,21 +73,38 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator.
+ *
+ * @param[in] dataset1 Dataset 1.
+ * @param[in] dataset2 Dataset 2.
+ */
iterator(const T_noref *dataset1, const U_noref *dataset2)
: _iter1{ dataset1->begin() }, _iter2{ dataset2->begin() }, _first_size{ dataset1->size() }
{
}
+ /** Get the description of the current value.
+ *
+ * @return description of the current value.
+ */
std::string description() const
{
return _first_size > 0 ? _iter1.description() : _iter2.description();
}
+ /** Get the value of the iterator.
+ *
+ * @return the value of the iterator.
+ */
JoinDataset::type operator*() const
{
return _first_size > 0 ? *_iter1 : *_iter2;
}
+ /** Inrement the iterator.
+ *
+ * @return *this;
+ */
iterator &operator++()
{
if(_first_size > 0)
diff --git a/tests/framework/datasets/RangeDataset.h b/tests/framework/datasets/RangeDataset.h
index 637abe0..a087566 100644
--- a/tests/framework/datasets/RangeDataset.h
+++ b/tests/framework/datasets/RangeDataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -60,6 +60,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
RangeDataset(RangeDataset &&) = default;
/** Type of the dataset. */
@@ -68,22 +69,40 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator.
+ *
+ * @param[in] name Dataset name.
+ * @param[in] start Dataset start value.
+ * @param[in] step Dataset step size.
+ */
iterator(std::string name, T start, T step)
: _name{ name }, _value{ start }, _step{ step }
{
}
+ /** Get the description of the current value.
+ *
+ * @return description of the current value.
+ */
std::string description() const
{
using support::cpp11::to_string;
return _name + "=" + to_string(_value);
}
+ /** Get the value of the iterator.
+ *
+ * @return the value of the iterator.
+ */
RangeDataset::type operator*() const
{
return std::make_tuple(_value);
}
+ /** Inrement the iterator.
+ *
+ * @return *this;
+ */
iterator &operator++()
{
_value += _step;
diff --git a/tests/framework/datasets/SingletonDataset.h b/tests/framework/datasets/SingletonDataset.h
index 1acb576..47a38ec 100644
--- a/tests/framework/datasets/SingletonDataset.h
+++ b/tests/framework/datasets/SingletonDataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -56,6 +56,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
SingletonDataset(SingletonDataset &&) = default;
/** Type of the dataset. */
@@ -64,29 +65,51 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator.
+ *
+ * @param[in] name Name of the dataset.
+ * @param[in] value The singleton value.
+ */
iterator(std::string name, const T *value)
: _name{ name }, _value{ value }
{
}
+ /** Default destructor. */
~iterator() = default;
+ /** Allow instances of this class to be copy constructed */
iterator(const iterator &) = default;
+ /** Allow instances of this class to be copied */
iterator &operator=(const iterator &) = default;
- iterator(iterator &&) = default;
+ /** Allow instances of this class to be move constructed */
+ iterator(iterator &&) = default;
+ /** Allow instances of this class to be moved */
iterator &operator=(iterator &&) = default;
+ /** Get the description of the current value.
+ *
+ * @return description of the current value.
+ */
std::string description() const
{
using support::cpp11::to_string;
return _name + "=" + to_string(*_value);
}
+ /** Get the value of the iterator.
+ *
+ * @return the value of the iterator.
+ */
SingletonDataset::type operator*() const
{
return std::make_tuple(*_value);
}
+ /** Inrement the iterator.
+ *
+ * @return *this;
+ */
iterator &operator++()
{
return *this;
diff --git a/tests/framework/datasets/ZipDataset.h b/tests/framework/datasets/ZipDataset.h
index b95b720..3d93b92 100644
--- a/tests/framework/datasets/ZipDataset.h
+++ b/tests/framework/datasets/ZipDataset.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -62,6 +62,7 @@
{
}
+ /** Allow instances of this class to be move constructed */
ZipDataset(ZipDataset &&) = default;
/** Type of the dataset. */
@@ -70,21 +71,38 @@
/** Iterator for the dataset. */
struct iterator
{
+ /** Construct an iterator.
+ *
+ * @param[in] iter1 Iterator 1.
+ * @param[in] iter2 Iterator 2.
+ */
iterator(iter1_type iter1, iter2_type iter2)
: _iter1{ std::move(iter1) }, _iter2{ std::move(iter2) }
{
}
+ /** Get the description of the current value.
+ *
+ * @return description of the current value.
+ */
std::string description() const
{
return _iter1.description() + ":" + _iter2.description();
}
+ /** Get the value of the iterator.
+ *
+ * @return the value of the iterator.
+ */
ZipDataset::type operator*() const
{
return std::tuple_cat(*_iter1, *_iter2);
}
+ /** Inrement the iterator.
+ *
+ * @return *this;
+ */
iterator &operator++()
{
++_iter1;
diff --git a/tests/framework/instruments/Instrument.h b/tests/framework/instruments/Instrument.h
index e25939a..ae4644b 100644
--- a/tests/framework/instruments/Instrument.h
+++ b/tests/framework/instruments/Instrument.h
@@ -57,27 +57,74 @@
template <typename T, ScaleFactor scale>
static std::unique_ptr<Instrument> make_instrument();
+ /** Default constructor. */
Instrument() = default;
+ /** Allow instances of this class to be copy constructed */
Instrument(const Instrument &) = default;
- Instrument(Instrument &&) = default;
+ /** Allow instances of this class to be move constructed */
+ Instrument(Instrument &&) = default;
+ /** Allow instances of this class to be copied */
Instrument &operator=(const Instrument &) = default;
+ /** Allow instances of this class to be moved */
Instrument &operator=(Instrument &&) = default;
- virtual ~Instrument() = default;
+ /** Default destructor. */
+ virtual ~Instrument() = default;
/** Identifier for the instrument */
virtual std::string id() const = 0;
- /** Start measuring. */
- virtual void start() = 0;
+ /** Start of the test
+ *
+ * Called before the test set up starts
+ */
+ virtual void test_start()
+ {
+ }
- /** Stop measuring. */
- virtual void stop() = 0;
+ /** Start measuring.
+ *
+ * Called just before the run of the test starts
+ */
+ virtual void start()
+ {
+ }
+ /** Stop measuring.
+ *
+ * Called just after the run of the test ends
+ */
+ virtual void stop()
+ {
+ }
+
+ /** End of the test
+ *
+ * Called after the test teardown ended
+ */
+ virtual void test_stop()
+ {
+ }
+ /** Map of measurements */
using MeasurementsMap = std::map<std::string, Measurement>;
- /** Return the latest measurement. */
- virtual MeasurementsMap measurements() const = 0;
+ /** Return the latest measurements.
+ *
+ * @return the latest measurements.
+ */
+ virtual MeasurementsMap measurements() const
+ {
+ return MeasurementsMap();
+ }
+
+ /** Return the latest test measurements.
+ *
+ * @return the latest test measurements.
+ */
+ virtual MeasurementsMap test_measurements() const
+ {
+ return MeasurementsMap();
+ }
protected:
std::string _unit{};
diff --git a/tests/framework/instruments/Instruments.cpp b/tests/framework/instruments/Instruments.cpp
index 64e87f9..6d65b01 100644
--- a/tests/framework/instruments/Instruments.cpp
+++ b/tests/framework/instruments/Instruments.cpp
@@ -59,6 +59,9 @@
{ "opencl_timer_us", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_US) },
{ "opencl_timer_ms", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_MS) },
{ "opencl_timer_s", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_TIMER, ScaleFactor::TIME_S) },
+ { "opencl_memory_usage", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::NONE) },
+ { "opencl_memory_usage_k", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::SCALE_1K) },
+ { "opencl_memory_usage_m", std::pair<InstrumentType, ScaleFactor>(InstrumentType::OPENCL_MEMORY_USAGE, ScaleFactor::SCALE_1M) },
};
try
diff --git a/tests/framework/instruments/Instruments.h b/tests/framework/instruments/Instruments.h
index fe4c719..705fc59 100644
--- a/tests/framework/instruments/Instruments.h
+++ b/tests/framework/instruments/Instruments.h
@@ -25,6 +25,7 @@
#define ARM_COMPUTE_TEST_INSTRUMENTS
#include "MaliCounter.h"
+#include "OpenCLMemoryUsage.h"
#include "OpenCLTimer.h"
#include "PMUCounter.h"
#include "SchedulerTimer.h"
@@ -50,6 +51,7 @@
MALI = 0x0300,
OPENCL_TIMER = 0x0400,
SCHEDULER_TIMER = 0x0500,
+ OPENCL_MEMORY_USAGE = 0x0600,
};
using InstrumentsDescription = std::pair<InstrumentType, ScaleFactor>;
@@ -157,6 +159,22 @@
throw std::invalid_argument("Unsupported instrument scale");
}
break;
+ case InstrumentType::OPENCL_MEMORY_USAGE:
+ switch(instrument.second)
+ {
+ case ScaleFactor::NONE:
+ stream << "OPENCL_MEMORY_USAGE";
+ break;
+ case ScaleFactor::SCALE_1K:
+ stream << "OPENCL_MEMORY_USAGE_K";
+ break;
+ case ScaleFactor::SCALE_1M:
+ stream << "OPENCL_MEMORY_USAGE_M";
+ break;
+ default:
+ throw std::invalid_argument("Unsupported instrument scale");
+ }
+ break;
case InstrumentType::ALL:
stream << "ALL";
break;
diff --git a/tests/framework/instruments/InstrumentsStats.cpp b/tests/framework/instruments/InstrumentsStats.cpp
index 6fad8f3..8f7d8a1 100644
--- a/tests/framework/instruments/InstrumentsStats.cpp
+++ b/tests/framework/instruments/InstrumentsStats.cpp
@@ -22,7 +22,7 @@
* SOFTWARE.
*/
#include "InstrumentsStats.h"
-#include "arm_compute/core/utils/misc/utility.h"
+#include "arm_compute/core/utils/misc/Utility.h"
namespace arm_compute
{
diff --git a/tests/framework/instruments/MaliCounter.h b/tests/framework/instruments/MaliCounter.h
index b7c3483..a3cc446 100644
--- a/tests/framework/instruments/MaliCounter.h
+++ b/tests/framework/instruments/MaliCounter.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -41,10 +41,15 @@
class MaliCounter : public Instrument
{
public:
- /** Default constructor. */
+ /** Default constructor.
+ *
+ * @param[in] scale_factor Measurement scale factor;
+ */
MaliCounter(ScaleFactor scale_factor);
+ /** Prevent instances of this class from being copy constructed */
MaliCounter(const MaliCounter &) = delete;
+ /** Prevent instances of this class from being copied */
MaliCounter &operator=(const MaliCounter &) = delete;
/** Default destructor */
diff --git a/tests/framework/instruments/Measurement.h b/tests/framework/instruments/Measurement.h
index 1beacf6..5c62977 100644
--- a/tests/framework/instruments/Measurement.h
+++ b/tests/framework/instruments/Measurement.h
@@ -40,6 +40,7 @@
/** Generic measurement that stores values as either double or long long int. */
struct Measurement
{
+ /** Measurement value */
struct Value
{
/** Constructor
@@ -187,6 +188,13 @@
}
}
+ /** Get the relative standard deviation to a given distribution as a percentage.
+ *
+ * @param[in] variance The variance of the distribution.
+ * @param[in] mean The mean of the distribution.
+ *
+ * @return the relative standard deviation.
+ */
static double relative_standard_deviation(const Value &variance, const Value &mean)
{
if(variance.is_floating_point)
@@ -222,6 +230,11 @@
/** Stream output operator to print the measurement.
*
* Prints value and unit.
+ *
+ * @param[out] os Output stream.
+ * @param[in] measurement Measurement.
+ *
+ * @return the modified output stream.
*/
friend inline std::ostream &operator<<(std::ostream &os, const Measurement &measurement)
{
diff --git a/tests/framework/instruments/OpenCLMemoryUsage.cpp b/tests/framework/instruments/OpenCLMemoryUsage.cpp
new file mode 100644
index 0000000..119ad4c
--- /dev/null
+++ b/tests/framework/instruments/OpenCLMemoryUsage.cpp
@@ -0,0 +1,208 @@
+/*
+ * Copyright (c) 2018 ARM Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#include "OpenCLMemoryUsage.h"
+
+#include "../Framework.h"
+#include "../Utils.h"
+
+#ifndef ARM_COMPUTE_CL
+#error "You can't use OpenCLMemoryUsage without OpenCL"
+#endif /* ARM_COMPUTE_CL */
+
+#include "arm_compute/core/CL/CLKernelLibrary.h"
+
+namespace arm_compute
+{
+namespace test
+{
+namespace framework
+{
+std::string OpenCLMemoryUsage::id() const
+{
+ return "OpenCLMemoryUsage";
+}
+
+OpenCLMemoryUsage::OpenCLMemoryUsage(ScaleFactor scale_factor)
+ : real_clCreateBuffer(CLSymbols::get().clCreateBuffer_ptr), real_clRetainMemObject(CLSymbols::get().clRetainMemObject_ptr), real_clReleaseMemObject(CLSymbols::get().clReleaseMemObject_ptr),
+ real_clSVMAlloc(CLSymbols::get().clSVMAlloc_ptr), real_clSVMFree(CLSymbols::get().clSVMFree_ptr), _allocations(), _svm_allocations(), _start(), _end(), _now()
+{
+ switch(scale_factor)
+ {
+ case ScaleFactor::NONE:
+ _scale_factor = 1;
+ _unit = "";
+ break;
+ case ScaleFactor::SCALE_1K:
+ _scale_factor = 1000;
+ _unit = "K ";
+ break;
+ case ScaleFactor::SCALE_1M:
+ _scale_factor = 1000000;
+ _unit = "M ";
+ break;
+ default:
+ ARM_COMPUTE_ERROR("Invalid scale");
+ }
+}
+
+void OpenCLMemoryUsage::test_start()
+{
+ _now = Stats();
+
+ ARM_COMPUTE_ERROR_ON(CLSymbols::get().clCreateBuffer_ptr == nullptr);
+ CLSymbols::get().clCreateBuffer_ptr = [this](
+ cl_context context,
+ cl_mem_flags flags,
+ size_t size,
+ void *host_ptr,
+ cl_int * errcode_ret)
+ {
+ cl_mem retval = this->real_clCreateBuffer(context, flags, size, host_ptr, errcode_ret);
+ if(host_ptr != nullptr)
+ {
+ // If it's an SVM / external allocation;
+ size = 0;
+ }
+ else
+ {
+ _now.num_allocations++;
+ _now.in_use += size;
+ _now.total_allocated += size;
+ if(_now.in_use > _now.max_in_use)
+ {
+ _now.max_in_use = _now.in_use;
+ }
+ }
+ this->_allocations[retval] = Allocation(size);
+ return retval;
+ };
+ ARM_COMPUTE_ERROR_ON(CLSymbols::get().clRetainMemObject_ptr == nullptr);
+ CLSymbols::get().clRetainMemObject_ptr = [this](cl_mem memobj)
+ {
+ cl_int retval = this->real_clRetainMemObject(memobj);
+ this->_allocations[memobj].refcount++;
+ return retval;
+ };
+ ARM_COMPUTE_ERROR_ON(CLSymbols::get().clReleaseMemObject_ptr == nullptr);
+ CLSymbols::get().clReleaseMemObject_ptr = [this](cl_mem memobj)
+ {
+ cl_int retval = this->real_clRetainMemObject(memobj);
+ Allocation &alloc = this->_allocations[memobj];
+ if(--alloc.refcount == 0)
+ {
+ _now.in_use -= alloc.size;
+ }
+ return retval;
+ };
+
+ //Only intercept the function if it exists:
+ if(CLSymbols::get().clSVMAlloc_ptr != nullptr)
+ {
+ CLSymbols::get().clSVMAlloc_ptr = [this](cl_context context, cl_svm_mem_flags flags, size_t size, cl_uint alignment)
+ {
+ void *retval = this->real_clSVMAlloc(context, flags, size, alignment);
+ if(retval != nullptr)
+ {
+ _svm_allocations[retval] = size;
+ _now.num_allocations++;
+ _now.in_use += size;
+ _now.total_allocated += size;
+ if(_now.in_use > _now.max_in_use)
+ {
+ _now.max_in_use = _now.in_use;
+ }
+ }
+ return retval;
+ };
+ }
+
+ //Only intercept the function if it exists:
+ if(CLSymbols::get().clSVMFree_ptr != nullptr)
+ {
+ CLSymbols::get().clSVMFree_ptr = [this](cl_context context, void *svm_pointer)
+ {
+ this->real_clSVMFree(context, svm_pointer);
+ auto iterator = _svm_allocations.find(svm_pointer);
+ if(iterator != _svm_allocations.end())
+ {
+ size_t size = iterator->second;
+ _svm_allocations.erase(iterator);
+ _now.in_use -= size;
+ }
+ };
+ }
+}
+
+void OpenCLMemoryUsage::start()
+{
+ _start = _now;
+}
+void OpenCLMemoryUsage::stop()
+{
+ _end = _now;
+}
+
+void OpenCLMemoryUsage::test_stop()
+{
+ // Restore real function
+ CLSymbols::get().clCreateBuffer_ptr = real_clCreateBuffer;
+ CLSymbols::get().clRetainMemObject_ptr = real_clRetainMemObject;
+ CLSymbols::get().clReleaseMemObject_ptr = real_clReleaseMemObject;
+ CLSymbols::get().clSVMAlloc_ptr = real_clSVMAlloc;
+ CLSymbols::get().clSVMFree_ptr = real_clSVMFree;
+}
+
+Instrument::MeasurementsMap OpenCLMemoryUsage::measurements() const
+{
+ MeasurementsMap measurements;
+ measurements.emplace("Num buffers allocated per run", Measurement(_end.num_allocations - _start.num_allocations, ""));
+ measurements.emplace("Total memory allocated per run", Measurement((_end.total_allocated - _start.total_allocated) / _scale_factor, _unit));
+ measurements.emplace("Memory in use at start of run", Measurement(_start.in_use / _scale_factor, _unit));
+
+ return measurements;
+}
+Instrument::MeasurementsMap OpenCLMemoryUsage::test_measurements() const
+{
+ MeasurementsMap measurements;
+ measurements.emplace("Num buffers", Measurement(_now.num_allocations, ""));
+ measurements.emplace("Total memory allocated", Measurement(_now.total_allocated / _scale_factor, _unit));
+ measurements.emplace("Max memory allocated", Measurement(_now.max_in_use / _scale_factor, _unit));
+ measurements.emplace("Memory leaked", Measurement(_now.in_use / _scale_factor, _unit));
+
+ size_t num_programs = CLKernelLibrary::get().get_built_programs().size();
+ size_t total_size = 0;
+ for(auto it : CLKernelLibrary::get().get_built_programs())
+ {
+ std::vector<size_t> binary_sizes = it.second.getInfo<CL_PROGRAM_BINARY_SIZES>();
+ total_size = std::accumulate(binary_sizes.begin(), binary_sizes.end(), total_size);
+ }
+
+ measurements.emplace("Num programs in cache", Measurement(num_programs, ""));
+ measurements.emplace("Total programs memory in cache", Measurement(total_size / _scale_factor, _unit));
+
+ return measurements;
+}
+} // namespace framework
+} // namespace test
+} // namespace arm_compute
diff --git a/tests/framework/instruments/OpenCLMemoryUsage.h b/tests/framework/instruments/OpenCLMemoryUsage.h
new file mode 100644
index 0000000..7593c01
--- /dev/null
+++ b/tests/framework/instruments/OpenCLMemoryUsage.h
@@ -0,0 +1,90 @@
+/*
+ * Copyright (c) 2018 ARM Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef ARM_COMPUTE_TEST_OPENCL_MEMORY_USAGE
+#define ARM_COMPUTE_TEST_OPENCL_MEMORY_USAGE
+
+#include "Instrument.h"
+
+#ifdef ARM_COMPUTE_CL
+#include "arm_compute/core/CL/OpenCL.h"
+#endif /* ARM_COMPUTE_CL */
+
+#include <list>
+
+namespace arm_compute
+{
+namespace test
+{
+namespace framework
+{
+/** Instrument collecting memory usage information for OpenCL*/
+class OpenCLMemoryUsage : public Instrument
+{
+public:
+ /** Construct an OpenCL timer.
+ *
+ * @param[in] scale_factor Measurement scale factor.
+ */
+ OpenCLMemoryUsage(ScaleFactor scale_factor);
+ std::string id() const override;
+ void test_start() override;
+ void start() override;
+ void stop() override;
+ void test_stop() override;
+ MeasurementsMap test_measurements() const override;
+ MeasurementsMap measurements() const override;
+#ifdef ARM_COMPUTE_CL
+ std::function<decltype(clCreateBuffer)> real_clCreateBuffer;
+ std::function<decltype(clRetainMemObject)> real_clRetainMemObject;
+ std::function<decltype(clReleaseMemObject)> real_clReleaseMemObject;
+ std::function<decltype(clSVMAlloc)> real_clSVMAlloc;
+ std::function<decltype(clSVMFree)> real_clSVMFree;
+
+private:
+ float _scale_factor{};
+ struct Allocation
+ {
+ Allocation() = default;
+ Allocation(size_t alloc_size)
+ : size(alloc_size)
+ {
+ }
+ size_t size{ 0 };
+ int refcount{ 1 };
+ };
+ std::map<cl_mem, Allocation> _allocations;
+ std::map<void *, size_t> _svm_allocations;
+ struct Stats
+ {
+ size_t total_allocated{ 0 };
+ size_t max_in_use{ 0 };
+ size_t in_use{ 0 };
+ size_t num_allocations{ 0 };
+ } _start, _end, _now;
+#endif /* ARM_COMPUTE_CL */
+};
+} // namespace framework
+} // namespace test
+} // namespace arm_compute
+#endif /* ARM_COMPUTE_TEST_OPENCL_MEMORY_USAGE */
diff --git a/tests/framework/instruments/OpenCLTimer.cpp b/tests/framework/instruments/OpenCLTimer.cpp
index 9743015..4391c43 100644
--- a/tests/framework/instruments/OpenCLTimer.cpp
+++ b/tests/framework/instruments/OpenCLTimer.cpp
@@ -26,6 +26,7 @@
#include "../Framework.h"
#include "../Utils.h"
+#include "arm_compute/graph/INode.h"
#include "arm_compute/runtime/CL/CLScheduler.h"
#ifndef ARM_COMPUTE_CL
@@ -43,55 +44,8 @@
return "OpenCLTimer";
}
-/* Function to be used to intercept kernel enqueues and store their OpenCL Event */
-class Interceptor
-{
-public:
- explicit Interceptor(OpenCLTimer &timer)
- : _timer(timer)
- {
- }
-
- cl_int operator()(
- cl_command_queue command_queue,
- cl_kernel kernel,
- cl_uint work_dim,
- const size_t *gwo,
- const size_t *gws,
- const size_t *lws,
- cl_uint num_events_in_wait_list,
- const cl_event *event_wait_list,
- cl_event *event)
- {
- ARM_COMPUTE_ERROR_ON_MSG(event != nullptr, "Not supported");
- ARM_COMPUTE_UNUSED(event);
-
- OpenCLTimer::kernel_info info;
- cl::Kernel cpp_kernel(kernel, true);
- std::stringstream ss;
- ss << cpp_kernel.getInfo<CL_KERNEL_FUNCTION_NAME>();
- if(gws != nullptr)
- {
- ss << " GWS[" << gws[0] << "," << gws[1] << "," << gws[2] << "]";
- }
- if(lws != nullptr)
- {
- ss << " LWS[" << lws[0] << "," << lws[1] << "," << lws[2] << "]";
- }
- info.name = ss.str();
- cl_event tmp;
- cl_int retval = _timer.real_function(command_queue, kernel, work_dim, gwo, gws, lws, num_events_in_wait_list, event_wait_list, &tmp);
- info.event = tmp;
- _timer.kernels.push_back(std::move(info));
- return retval;
- }
-
-private:
- OpenCLTimer &_timer;
-};
-
OpenCLTimer::OpenCLTimer(ScaleFactor scale_factor)
- : real_function(CLSymbols::get().clEnqueueNDRangeKernel_ptr)
+ : _kernels(), _real_function(nullptr), _real_graph_function(nullptr), _prefix(), _timer_enabled(false)
{
auto q = CLScheduler::get().queue();
cl_command_queue_properties props = q.getInfo<CL_QUEUE_PROPERTIES>();
@@ -123,24 +77,97 @@
}
}
-void OpenCLTimer::start()
+void OpenCLTimer::test_start()
{
- kernels.clear();
// Start intercepting enqueues:
- CLSymbols::get().clEnqueueNDRangeKernel_ptr = Interceptor(*this);
+ ARM_COMPUTE_ERROR_ON(_real_function != nullptr);
+ ARM_COMPUTE_ERROR_ON(_real_graph_function != nullptr);
+ _real_function = CLSymbols::get().clEnqueueNDRangeKernel_ptr;
+ _real_graph_function = graph::TaskExecutor::get().execute_function;
+ auto interceptor = [this](
+ cl_command_queue command_queue,
+ cl_kernel kernel,
+ cl_uint work_dim,
+ const size_t *gwo,
+ const size_t *gws,
+ const size_t *lws,
+ cl_uint num_events_in_wait_list,
+ const cl_event * event_wait_list,
+ cl_event * event)
+ {
+ if(this->_timer_enabled)
+ {
+ ARM_COMPUTE_ERROR_ON_MSG(event != nullptr, "Not supported");
+ ARM_COMPUTE_UNUSED(event);
+
+ OpenCLTimer::kernel_info info;
+ cl::Kernel cpp_kernel(kernel, true);
+ std::stringstream ss;
+ ss << this->_prefix << cpp_kernel.getInfo<CL_KERNEL_FUNCTION_NAME>();
+ if(gws != nullptr)
+ {
+ ss << " GWS[" << gws[0] << "," << gws[1] << "," << gws[2] << "]";
+ }
+ if(lws != nullptr)
+ {
+ ss << " LWS[" << lws[0] << "," << lws[1] << "," << lws[2] << "]";
+ }
+ info.name = ss.str();
+ cl_event tmp;
+ cl_int retval = this->_real_function(command_queue, kernel, work_dim, gwo, gws, lws, num_events_in_wait_list, event_wait_list, &tmp);
+ info.event = tmp;
+ this->_kernels.push_back(std::move(info));
+ return retval;
+ }
+ else
+ {
+ return this->_real_function(command_queue, kernel, work_dim, gwo, gws, lws, num_events_in_wait_list, event_wait_list, event);
+ }
+ };
+
+ // Start intercepting tasks:
+ auto task_interceptor = [this](graph::ExecutionTask & task)
+ {
+ if(task.node != nullptr && !task.node->name().empty())
+ {
+ this->_prefix = task.node->name() + "/";
+ }
+ else
+ {
+ this->_prefix = "";
+ }
+ this->_real_graph_function(task);
+ this->_prefix = "";
+ };
+
+ CLSymbols::get().clEnqueueNDRangeKernel_ptr = interceptor;
+ graph::TaskExecutor::get().execute_function = task_interceptor;
}
+void OpenCLTimer::start()
+{
+ _kernels.clear();
+ _timer_enabled = true;
+}
void OpenCLTimer::stop()
{
+ _timer_enabled = false;
+}
+
+void OpenCLTimer::test_stop()
+{
// Restore real function
- CLSymbols::get().clEnqueueNDRangeKernel_ptr = real_function;
+ CLSymbols::get().clEnqueueNDRangeKernel_ptr = _real_function;
+ graph::TaskExecutor::get().execute_function = _real_graph_function;
+ _real_graph_function = nullptr;
+ _real_function = nullptr;
}
Instrument::MeasurementsMap OpenCLTimer::measurements() const
{
MeasurementsMap measurements;
unsigned int kernel_number = 0;
- for(auto kernel : kernels)
+ for(auto kernel : _kernels)
{
cl_ulong start = kernel.event.getProfilingInfo<CL_PROFILING_COMMAND_START>();
cl_ulong end = kernel.event.getProfilingInfo<CL_PROFILING_COMMAND_END>();
diff --git a/tests/framework/instruments/OpenCLTimer.h b/tests/framework/instruments/OpenCLTimer.h
index a3dc107..c5f3bce 100644
--- a/tests/framework/instruments/OpenCLTimer.h
+++ b/tests/framework/instruments/OpenCLTimer.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -30,6 +30,8 @@
#include "arm_compute/core/CL/OpenCL.h"
#endif /* ARM_COMPUTE_CL */
+#include "arm_compute/graph/Workload.h"
+
#include <list>
namespace arm_compute
@@ -42,19 +44,30 @@
class OpenCLTimer : public Instrument
{
public:
+ /** Construct an OpenCL timer.
+ *
+ * @param[in] scale_factor Measurement scale factor.
+ */
OpenCLTimer(ScaleFactor scale_factor);
std::string id() const override;
+ void test_start() override;
void start() override;
void stop() override;
+ void test_stop() override;
MeasurementsMap measurements() const override;
+
+private:
#ifdef ARM_COMPUTE_CL
struct kernel_info
{
cl::Event event{}; /**< OpenCL event associated to the kernel enqueue */
std::string name{}; /**< OpenCL Kernel name */
};
- std::list<kernel_info> kernels{};
- std::function<decltype(clEnqueueNDRangeKernel)> real_function;
+ std::list<kernel_info> _kernels;
+ std::function<decltype(clEnqueueNDRangeKernel)> _real_function;
+ std::function<decltype(graph::execute_task)> _real_graph_function;
+ std::string _prefix;
+ bool _timer_enabled;
#endif /* ARM_COMPUTE_CL */
private:
diff --git a/tests/framework/instruments/PMU.h b/tests/framework/instruments/PMU.h
index c069a63..1dc41be 100644
--- a/tests/framework/instruments/PMU.h
+++ b/tests/framework/instruments/PMU.h
@@ -64,10 +64,16 @@
template <typename T>
T get_value() const;
- /** Open the specified counter based on the default configuration. */
+ /** Open the specified counter based on the default configuration.
+ *
+ * @param[in] config The default configuration.
+ */
void open(uint64_t config);
- /** Open the specified configuration. */
+ /** Open the specified configuration.
+ *
+ * @param[in] perf_config The specified configuration.
+ */
void open(const perf_event_attr &perf_config);
/** Close the currently open counter. */
diff --git a/tests/framework/instruments/PMUCounter.h b/tests/framework/instruments/PMUCounter.h
index e1b9433..0719b10 100644
--- a/tests/framework/instruments/PMUCounter.h
+++ b/tests/framework/instruments/PMUCounter.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -37,6 +37,10 @@
class PMUCounter : public Instrument
{
public:
+ /** Construct a PMU counter.
+ *
+ * @param[in] scale_factor Measurement scale factor.
+ */
PMUCounter(ScaleFactor scale_factor)
{
switch(scale_factor)
diff --git a/tests/framework/instruments/SchedulerTimer.cpp b/tests/framework/instruments/SchedulerTimer.cpp
index e42cebd..1b37b18 100644
--- a/tests/framework/instruments/SchedulerTimer.cpp
+++ b/tests/framework/instruments/SchedulerTimer.cpp
@@ -25,6 +25,8 @@
#include "WallClockTimer.h"
#include "arm_compute/core/CPP/ICPPKernel.h"
+#include "arm_compute/core/utils/misc/Cast.h"
+#include "arm_compute/graph/INode.h"
namespace arm_compute
{
@@ -42,7 +44,7 @@
public:
/** Default constructor. */
Interceptor(std::list<SchedulerTimer::kernel_info> &kernels, IScheduler &real_scheduler, ScaleFactor scale_factor)
- : _kernels(kernels), _real_scheduler(real_scheduler), _timer(scale_factor)
+ : _kernels(kernels), _real_scheduler(real_scheduler), _timer(scale_factor), _prefix()
{
}
@@ -56,6 +58,11 @@
return _real_scheduler.num_threads();
}
+ void set_prefix(std::string prefix)
+ {
+ _prefix = std::move(prefix);
+ }
+
void schedule(ICPPKernel *kernel, unsigned int split_dimension) override
{
_timer.start();
@@ -64,6 +71,7 @@
SchedulerTimer::kernel_info info;
info.name = kernel->name();
+ info.prefix = _prefix;
info.measurements = _timer.measurements();
_kernels.push_back(std::move(info));
}
@@ -72,32 +80,68 @@
std::list<SchedulerTimer::kernel_info> &_kernels;
IScheduler &_real_scheduler;
WallClockTimer _timer;
+ std::string _prefix;
};
SchedulerTimer::SchedulerTimer(ScaleFactor scale_factor)
- : _kernels(), _real_scheduler(nullptr), _real_scheduler_type(), _scale_factor(scale_factor)
+ : _kernels(), _real_scheduler(nullptr), _real_scheduler_type(), _real_graph_function(nullptr), _scale_factor(scale_factor), _interceptor(nullptr)
{
}
-void SchedulerTimer::start()
+void SchedulerTimer::test_start()
{
+ // Start intercepting tasks:
+ ARM_COMPUTE_ERROR_ON(_real_graph_function != nullptr);
+ _real_graph_function = graph::TaskExecutor::get().execute_function;
+ auto task_interceptor = [this](graph::ExecutionTask & task)
+ {
+ Interceptor *scheduler = nullptr;
+ if(dynamic_cast<Interceptor *>(this->_interceptor.get()) != nullptr)
+ {
+ scheduler = arm_compute::utils::cast::polymorphic_downcast<Interceptor *>(_interceptor.get());
+ if(task.node != nullptr && !task.node->name().empty())
+ {
+ scheduler->set_prefix(task.node->name() + "/");
+ }
+ else
+ {
+ scheduler->set_prefix("");
+ }
+ }
+
+ this->_real_graph_function(task);
+
+ if(scheduler != nullptr)
+ {
+ scheduler->set_prefix("");
+ }
+ };
+
ARM_COMPUTE_ERROR_ON(_real_scheduler != nullptr);
_real_scheduler_type = Scheduler::get_type();
//Note: We can't currently replace a custom scheduler
if(_real_scheduler_type != Scheduler::Type::CUSTOM)
{
- _real_scheduler = &Scheduler::get();
- auto interceptor = std::make_shared<Interceptor>(_kernels, *_real_scheduler, _scale_factor);
- Scheduler::set(std::static_pointer_cast<IScheduler>(interceptor));
+ _real_scheduler = &Scheduler::get();
+ _interceptor = std::make_shared<Interceptor>(_kernels, *_real_scheduler, _scale_factor);
+ Scheduler::set(std::static_pointer_cast<IScheduler>(_interceptor));
+ graph::TaskExecutor::get().execute_function = task_interceptor;
}
+}
+
+void SchedulerTimer::start()
+{
_kernels.clear();
}
-void SchedulerTimer::stop()
+void SchedulerTimer::test_stop()
{
// Restore real scheduler
Scheduler::set(_real_scheduler_type);
- _real_scheduler = nullptr;
+ _real_scheduler = nullptr;
+ _interceptor = nullptr;
+ graph::TaskExecutor::get().execute_function = _real_graph_function;
+ _real_graph_function = nullptr;
}
Instrument::MeasurementsMap SchedulerTimer::measurements() const
@@ -106,7 +150,7 @@
unsigned int kernel_number = 0;
for(auto kernel : _kernels)
{
- measurements.emplace(kernel.name + " #" + support::cpp11::to_string(kernel_number++), kernel.measurements.begin()->second);
+ measurements.emplace(kernel.prefix + kernel.name + " #" + support::cpp11::to_string(kernel_number++), kernel.measurements.begin()->second);
}
return measurements;
diff --git a/tests/framework/instruments/SchedulerTimer.h b/tests/framework/instruments/SchedulerTimer.h
index 446506a..55d5f25 100644
--- a/tests/framework/instruments/SchedulerTimer.h
+++ b/tests/framework/instruments/SchedulerTimer.h
@@ -25,7 +25,9 @@
#define ARM_COMPUTE_TEST_SCHEDULER_TIMER
#include "Instrument.h"
+#include "arm_compute/graph/Workload.h"
#include "arm_compute/runtime/Scheduler.h"
+
#include <list>
namespace arm_compute
@@ -38,24 +40,38 @@
class SchedulerTimer : public Instrument
{
public:
- SchedulerTimer(const SchedulerTimer &) = delete;
- SchedulerTimer &operator=(const SchedulerTimer &) = delete;
+ /** Construct a Scheduler timer.
+ *
+ * @param[in] scale_factor Measurement scale factor.
+ */
SchedulerTimer(ScaleFactor scale_factor);
+
+ /** Prevent instances of this class from being copy constructed */
+ SchedulerTimer(const SchedulerTimer &) = delete;
+ /** Prevent instances of this class from being copied */
+ SchedulerTimer &operator=(const SchedulerTimer &) = delete;
+
std::string id() const override;
+ void test_start() override;
void start() override;
- void stop() override;
+ void test_stop() override;
Instrument::MeasurementsMap measurements() const override;
+
+ /** Kernel information */
struct kernel_info
{
Instrument::MeasurementsMap measurements{}; /**< Time it took the kernel to run */
std::string name{}; /**< Kernel name */
+ std::string prefix{}; /**< Kernel prefix */
};
private:
- std::list<kernel_info> _kernels;
- IScheduler *_real_scheduler;
- Scheduler::Type _real_scheduler_type;
- ScaleFactor _scale_factor;
+ std::list<kernel_info> _kernels;
+ IScheduler *_real_scheduler;
+ Scheduler::Type _real_scheduler_type;
+ std::function<decltype(graph::execute_task)> _real_graph_function;
+ ScaleFactor _scale_factor;
+ std::shared_ptr<IScheduler> _interceptor;
};
} // namespace framework
} // namespace test
diff --git a/tests/framework/instruments/WallClockTimer.h b/tests/framework/instruments/WallClockTimer.h
index 468f4d3..c9829ae 100644
--- a/tests/framework/instruments/WallClockTimer.h
+++ b/tests/framework/instruments/WallClockTimer.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -38,6 +38,10 @@
class WallClockTimer : public Instrument
{
public:
+ /** Construct a Wall clock timer.
+ *
+ * @param[in] scale_factor Measurement scale factor.
+ */
WallClockTimer(ScaleFactor scale_factor)
{
switch(scale_factor)
diff --git a/tests/framework/instruments/hwc.hpp b/tests/framework/instruments/hwc.hpp
index 3607ef5..8c48e0c 100644
--- a/tests/framework/instruments/hwc.hpp
+++ b/tests/framework/instruments/hwc.hpp
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -37,6 +37,8 @@
#include <sys/mman.h>
#include <unistd.h>
+#ifndef DOXYGEN_SKIP_THIS
+
#if defined(ANDROID) || defined(__ANDROID__)
/* We use _IOR_BAD/_IOW_BAD rather than _IOR/_IOW otherwise fails to compile with NDK-BUILD because of _IOC_TYPECHECK is defined, not because the paramter is invalid */
#define MALI_IOR(a, b, c) _IOR_BAD(a, b, c)
@@ -387,4 +389,7 @@
return 0;
}
} // namespace mali_userspace
+
+#endif /* DOXYGEN_SKIP_THIS */
+
#endif /* ARM_COMPUTE_TEST_HWC */
diff --git a/tests/framework/instruments/hwc_names.hpp b/tests/framework/instruments/hwc_names.hpp
index ffc19b5..cbcb0e7 100644
--- a/tests/framework/instruments/hwc_names.hpp
+++ b/tests/framework/instruments/hwc_names.hpp
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -24,6 +24,8 @@
#ifndef ARM_COMPUTE_TEST_HWC_NAMES
#define ARM_COMPUTE_TEST_HWC_NAMES
+#ifndef DOXYGEN_SKIP_THIS
+
namespace mali_userspace
{
enum MaliCounterBlockName
@@ -3056,4 +3058,7 @@
NUM_PRODUCTS = sizeof(products) / sizeof(products[0])
};
} // namespace mali_userspace
+
+#endif /* DOXYGEN_SKIP_THIS */
+
#endif /* ARM_COMPUTE_TEST_HWC_NAMES */
diff --git a/tests/framework/printers/PrettyPrinter.cpp b/tests/framework/printers/PrettyPrinter.cpp
index ef8f91a..3181951 100644
--- a/tests/framework/printers/PrettyPrinter.cpp
+++ b/tests/framework/printers/PrettyPrinter.cpp
@@ -129,8 +129,8 @@
if(instrument.second.size() > 1)
{
*_stream << ", STDDEV=" << arithmetic_to_string(stats.relative_standard_deviation(), 2) << " %";
- *_stream << ", MIN=" << stats.min() << ", ";
- *_stream << ", MAX=" << stats.max() << ", ";
+ *_stream << ", MIN=" << stats.min();
+ *_stream << ", MAX=" << stats.max();
*_stream << ", MEDIAN=" << stats.median().value() << " " << stats.median().unit();
}
*_stream << end_color() << "\n";
diff --git a/tests/framework/printers/Printer.h b/tests/framework/printers/Printer.h
index cb0aa1e..cbe22fb 100644
--- a/tests/framework/printers/Printer.h
+++ b/tests/framework/printers/Printer.h
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -55,11 +55,16 @@
*/
Printer(std::ostream &stream);
+ /** Prevent instances of this class from being copy constructed */
Printer(const Printer &) = delete;
+ /** Prevent instances of this class from being copied */
Printer &operator=(const Printer &) = delete;
- Printer(Printer &&) = default;
+ /** Allow instances of this class to be move constructed */
+ Printer(Printer &&) = default;
+ /** Allow instances of this class to be moved */
Printer &operator=(Printer &&) = default;
+ /** Default destructor. */
virtual ~Printer() = default;
/** Print given string.