aboutsummaryrefslogtreecommitdiff
path: root/src
diff options
context:
space:
mode:
authorGeorgios Pinitas <georgios.pinitas@arm.com>2017-12-22 15:27:52 +0000
committerAnthony Barbier <anthony.barbier@arm.com>2018-11-02 16:49:16 +0000
commitd8734b55d89f05901ba9a75349761a9c955d9243 (patch)
treee23d53a0fb73251f7416993e4d3a7241e533e79e /src
parent7390e05561a5c49306ebbf2eb2dcb1848546f201 (diff)
downloadComputeLibrary-d8734b55d89f05901ba9a75349761a9c955d9243.tar.gz
COMPMID-793 : Add graph intermediate representation
Change-Id: Ic1685de4e19e0ac79669ef2da64e1dc96c7ea0bf Reviewed-on: https://eu-gerrit-1.euhpc.arm.com/115248 Tested-by: Jenkins <bsgcomp@arm.com> Reviewed-by: Anthony Barbier <anthony.barbier@arm.com>
Diffstat (limited to 'src')
-rw-r--r--src/core/NEON/kernels/NEIm2ColKernel.cpp3
-rw-r--r--src/core/NEON/kernels/NEMinMaxLocationKernel.cpp4
-rw-r--r--src/core/NEON/kernels/NESoftmaxLayerKernel.cpp2
-rw-r--r--src/graph2/Graph.cpp227
-rw-r--r--src/graph2/GraphBuilder.cpp373
-rw-r--r--src/graph2/GraphContext.cpp84
-rw-r--r--src/graph2/GraphManager.cpp108
-rw-r--r--src/graph2/INode.cpp193
-rw-r--r--src/graph2/PassManager.cpp88
-rw-r--r--src/graph2/Tensor.cpp111
-rw-r--r--src/graph2/Utils.cpp97
-rw-r--r--src/graph2/Workload.cpp41
-rw-r--r--src/graph2/backends/BackendRegistry.cpp63
-rw-r--r--src/graph2/backends/CL/CLDeviceBackend.cpp175
-rw-r--r--src/graph2/backends/CL/CLFunctionsFactory.cpp584
-rw-r--r--src/graph2/backends/CL/CLSubTensorHandle.cpp73
-rw-r--r--src/graph2/backends/CL/CLTensorHandle.cpp69
-rw-r--r--src/graph2/backends/NEON/NEDeviceBackend.cpp135
-rw-r--r--src/graph2/backends/NEON/NEFunctionFactory.cpp563
-rw-r--r--src/graph2/backends/NEON/NESubTensorHandle.cpp70
-rw-r--r--src/graph2/backends/NEON/NETensorHandle.cpp68
-rw-r--r--src/graph2/detail/ExecutionHelpers.cpp170
-rw-r--r--src/graph2/frontend/Stream.cpp70
-rw-r--r--src/graph2/frontend/SubStream.cpp59
-rw-r--r--src/graph2/mutators/DepthConcatSubTensorMutator.cpp86
-rw-r--r--src/graph2/mutators/InPlaceOperationMutator.cpp63
-rw-r--r--src/graph2/mutators/NodeFusionMutator.cpp96
-rw-r--r--src/graph2/nodes/ActivationLayerNode.cpp83
-rw-r--r--src/graph2/nodes/BatchNormalizationLayerNode.cpp94
-rw-r--r--src/graph2/nodes/ConstNode.cpp72
-rw-r--r--src/graph2/nodes/ConvolutionLayerNode.cpp111
-rw-r--r--src/graph2/nodes/DepthConcatenateLayerNode.cpp133
-rw-r--r--src/graph2/nodes/DepthwiseConvolutionLayerNode.cpp110
-rw-r--r--src/graph2/nodes/EltwiseLayerNode.cpp83
-rw-r--r--src/graph2/nodes/FlattenLayerNode.cpp80
-rw-r--r--src/graph2/nodes/FullyConnectedLayer.cpp107
-rw-r--r--src/graph2/nodes/InputNode.cpp72
-rw-r--r--src/graph2/nodes/NormalizationLayerNode.cpp84
-rw-r--r--src/graph2/nodes/OutputNode.cpp66
-rw-r--r--src/graph2/nodes/PoolingLayerNode.cpp103
-rw-r--r--src/graph2/nodes/ReshapeLayer.cpp81
-rw-r--r--src/graph2/nodes/SoftmaxLayerNode.cpp84
-rw-r--r--src/graph2/printers/DotGraphPrinter.cpp173
-rw-r--r--src/runtime/CL/CLSubTensor.cpp5
-rw-r--r--src/runtime/CL/functions/CLGEMM.cpp6
-rw-r--r--src/runtime/CL/functions/CLGEMMConvolutionLayer.cpp10
-rw-r--r--src/runtime/NEON/functions/NEConvolutionLayer.cpp1
-rw-r--r--src/runtime/SubTensor.cpp5
48 files changed, 5224 insertions, 14 deletions
diff --git a/src/core/NEON/kernels/NEIm2ColKernel.cpp b/src/core/NEON/kernels/NEIm2ColKernel.cpp
index 4fa329bf44..dee1608c43 100644
--- a/src/core/NEON/kernels/NEIm2ColKernel.cpp
+++ b/src/core/NEON/kernels/NEIm2ColKernel.cpp
@@ -319,8 +319,7 @@ void NEIm2ColKernel::configure(const ITensor *input, ITensor *output, const Size
ARM_COMPUTE_ERROR_ON_NULLPTR(input, output);
// Perform validation step
- ARM_COMPUTE_UNUSED(is_fully_connected);
- ARM_COMPUTE_UNUSED(is_flatten);
+ ARM_COMPUTE_UNUSED(is_fully_connected, is_flatten);
ARM_COMPUTE_ERROR_THROW_ON(validate_arguments(input->info(), output->info(), kernel_dims, conv_info, has_bias, is_fully_connected, is_flatten));
_input = input;
diff --git a/src/core/NEON/kernels/NEMinMaxLocationKernel.cpp b/src/core/NEON/kernels/NEMinMaxLocationKernel.cpp
index ad66acd8fa..b90e81339b 100644
--- a/src/core/NEON/kernels/NEMinMaxLocationKernel.cpp
+++ b/src/core/NEON/kernels/NEMinMaxLocationKernel.cpp
@@ -1,5 +1,5 @@
/*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
*
* SPDX-License-Identifier: MIT
*
@@ -32,7 +32,7 @@
#include "arm_compute/core/Types.h"
#include "arm_compute/core/Validate.h"
#include "arm_compute/core/Window.h"
-#include "arm_compute/core/utils/misc/utility.h"
+#include "arm_compute/core/utils/misc/Utility.h"
#include <algorithm>
#include <arm_neon.h>
diff --git a/src/core/NEON/kernels/NESoftmaxLayerKernel.cpp b/src/core/NEON/kernels/NESoftmaxLayerKernel.cpp
index 13d87a0989..d91efd267f 100644
--- a/src/core/NEON/kernels/NESoftmaxLayerKernel.cpp
+++ b/src/core/NEON/kernels/NESoftmaxLayerKernel.cpp
@@ -33,7 +33,7 @@
#include "arm_compute/core/Utils.h"
#include "arm_compute/core/Validate.h"
#include "arm_compute/core/Window.h"
-#include "arm_compute/core/utils/misc/utility.h"
+#include "arm_compute/core/utils/misc/Utility.h"
#include <algorithm>
#include <arm_neon.h>
diff --git a/src/graph2/Graph.cpp b/src/graph2/Graph.cpp
new file mode 100644
index 0000000000..ead67bc85a
--- /dev/null
+++ b/src/graph2/Graph.cpp
@@ -0,0 +1,227 @@
+/*
+ * 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 "arm_compute/graph2/Graph.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+Graph::Graph(GraphID id, std::string name)
+ : _id(id), _name(std::move(name)), _nodes(), _edges(), _tensors(), _tagged_nodes(), _mtx()
+{
+}
+
+bool Graph::remove_node(NodeID nid)
+{
+ if(nid >= _nodes.size())
+ {
+ return false;
+ }
+
+ std::unique_ptr<INode> &node = _nodes[nid];
+
+ // Remove node connections
+ if(node)
+ {
+ for(auto &input_eid : node->_input_edges)
+ {
+ remove_connection(input_eid);
+ }
+ for(auto &outpud_eid : node->_output_edges)
+ {
+ remove_connection(outpud_eid);
+ }
+ }
+
+ node = nullptr;
+
+ return true;
+}
+
+EdgeID Graph::add_connection(NodeID source, size_t source_idx, NodeID sink, size_t sink_idx)
+{
+ std::lock_guard<arm_compute::Mutex> lock(_mtx);
+
+ // Check if node index is valid, if node exists and finally if the connection index is valid
+ ARM_COMPUTE_ERROR_ON((source >= _nodes.size()) || (_nodes[source] == nullptr) || (source_idx >= _nodes[source]->num_outputs()));
+ ARM_COMPUTE_ERROR_ON((sink >= _nodes.size()) || (_nodes[sink] == nullptr) || (sink_idx >= _nodes[sink]->num_inputs()));
+
+ // Get nodes
+ std::unique_ptr<INode> &source_node = _nodes[source];
+ std::unique_ptr<INode> &sink_node = _nodes[sink];
+
+ // Check for duplicate connections (Check only sink node)
+ Edge *sink_node_edge = sink_node->input_edge(sink_idx);
+ if((sink_node_edge != nullptr) && (sink_node_edge->producer_id() == source) && (sink_node_edge->producer_idx() == source_idx)
+ && (sink_node_edge->consumer_id() == sink) && (sink_node_edge->consumer_idx() == sink_idx))
+ {
+ return sink_node_edge->id();
+ }
+
+ // Check if there is already a tensor associated with output if not create one
+ TensorID tid = source_node->output_id(source_idx);
+ if(tid == NullTensorID)
+ {
+ tid = create_tensor();
+ }
+ std::unique_ptr<Tensor> &tensor = _tensors[tid];
+
+ // Create connections
+ EdgeID eid = _edges.size();
+ auto connection = arm_compute::support::cpp14::make_unique<Edge>(eid, source_node.get(), source_idx, sink_node.get(), sink_idx, tensor.get());
+ _edges.push_back(std::move(connection));
+
+ // Add connections to source and sink nodes
+ source_node->_output_edges.insert(eid);
+ sink_node->_input_edges[sink_idx] = eid;
+
+ // Set tensor output node
+ source_node->_outputs[source_idx] = tid;
+
+ // Bind tensor to the edge
+ tensor->bind_edge(eid);
+
+ // Try and propagate shapes in sink node
+ sink_node->forward_descriptors();
+
+ return eid;
+}
+
+bool Graph::remove_connection(EdgeID eid)
+{
+ if(eid >= _edges.size())
+ {
+ return false;
+ }
+
+ std::unique_ptr<Edge> &edge = _edges[eid];
+
+ // Remove node connections
+ if(edge != nullptr)
+ {
+ // Get tensor bound to the edge
+ if(edge->tensor() != nullptr)
+ {
+ edge->tensor()->unbind_edge(eid);
+ }
+
+ // Remove edges from source node
+ if(edge->producer() != nullptr)
+ {
+ edge->producer()->_output_edges.erase(eid);
+ }
+
+ // Remove edges from sink node
+ if((edge->consumer() != nullptr) && (edge->consumer_idx() < edge->consumer()->_input_edges.size()))
+ {
+ edge->consumer()->_input_edges[edge->consumer_idx()] = EmptyEdgeID;
+ }
+ }
+
+ // Clear edge
+ edge = nullptr;
+
+ return true;
+}
+
+TensorID Graph::create_tensor(TensorDescriptor desc)
+{
+ TensorID tid = _tensors.size();
+ auto tensor = support::cpp14::make_unique<Tensor>(tid, desc);
+ _tensors.push_back(std::move(tensor));
+
+ return tid;
+}
+
+std::string Graph::name() const
+{
+ return _name;
+}
+
+GraphID Graph::id() const
+{
+ return _id;
+}
+
+const std::vector<NodeID> &Graph::inputs()
+{
+ return _tagged_nodes[NodeType::Input];
+}
+
+std::vector<std::unique_ptr<INode>> &Graph::nodes()
+{
+ return _nodes;
+}
+
+const std::vector<std::unique_ptr<INode>> &Graph::nodes() const
+{
+ return _nodes;
+}
+
+const std::vector<std::unique_ptr<Edge>> &Graph::edges() const
+{
+ return _edges;
+}
+
+std::vector<std::unique_ptr<Tensor>> &Graph::tensors()
+{
+ return _tensors;
+}
+
+const std::vector<std::unique_ptr<Tensor>> &Graph::tensors() const
+{
+ return _tensors;
+}
+
+const INode *Graph::node(NodeID id) const
+{
+ return (id >= _nodes.size()) ? nullptr : _nodes[id].get();
+}
+
+INode *Graph::node(NodeID id)
+{
+ return (id >= _nodes.size()) ? nullptr : _nodes[id].get();
+}
+
+const Edge *Graph::edge(EdgeID id) const
+{
+ return (id >= _edges.size()) ? nullptr : _edges[id].get();
+}
+
+Edge *Graph::edge(EdgeID id)
+{
+ return (id >= _edges.size()) ? nullptr : _edges[id].get();
+}
+
+const Tensor *Graph::tensor(TensorID id) const
+{
+ return (id >= _tensors.size()) ? nullptr : _tensors[id].get();
+}
+
+Tensor *Graph::tensor(TensorID id)
+{
+ return (id >= _tensors.size()) ? nullptr : _tensors[id].get();
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/GraphBuilder.cpp b/src/graph2/GraphBuilder.cpp
new file mode 100644
index 0000000000..aaf70c4e61
--- /dev/null
+++ b/src/graph2/GraphBuilder.cpp
@@ -0,0 +1,373 @@
+/*
+ * 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 "arm_compute/graph2/GraphBuilder.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Utils.h"
+#include "arm_compute/graph2/algorithms/BFS.h"
+#include "arm_compute/graph2/nodes/Nodes.h"
+
+#define CHECK_NODEIDX_PAIR(pair, g) \
+ ARM_COMPUTE_ERROR_ON(((pair).node_id >= (g).nodes().size()) || ((g).node((pair).node_id) == nullptr) || ((pair).index >= (g).node((pair).node_id)->num_outputs()));
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace
+{
+Status set_node_params(Graph &g, NodeID nid, NodeParams &params)
+{
+ INode *node = g.node(nid);
+ ARM_COMPUTE_RETURN_ERROR_ON(!node);
+
+ node->set_common_node_parameters(params);
+
+ return Status{};
+}
+Status set_accessor_on_node(Graph &g, NodeID nid, bool is_output, size_t idx, ITensorAccessorUPtr accessor)
+{
+ INode *node = g.node(nid);
+ ARM_COMPUTE_RETURN_ERROR_ON(!node);
+
+ Tensor *tensor = is_output ? node->output(idx) : node->input(idx);
+ ARM_COMPUTE_RETURN_ERROR_ON(!tensor);
+
+ tensor->set_accessor(std::move(accessor));
+
+ return Status{};
+}
+
+NodeID add_const_node_with_name(Graph &g, NodeParams params, const std::string &name, TensorDescriptor desc, ITensorAccessorUPtr accessor)
+{
+ params.name = params.name.empty() ? "" : params.name + name;
+ auto nid = GraphBuilder::add_const_node(g, params, desc, std::move(accessor));
+ set_node_params(g, nid, params);
+ return nid;
+}
+} // namespace
+
+NodeID GraphBuilder::add_const_node(Graph &g, NodeParams params, TensorDescriptor desc, ITensorAccessorUPtr accessor)
+{
+ auto nid = g.add_node<ConstNode>(desc);
+ set_node_params(g, nid, params);
+ set_accessor_on_node(g, nid, true, 0, std::move(accessor));
+ return nid;
+}
+
+NodeID GraphBuilder::add_input_node(Graph &g, NodeParams params, TensorDescriptor desc, ITensorAccessorUPtr accessor)
+{
+ auto nid = g.add_node<InputNode>(desc);
+ set_node_params(g, nid, params);
+ set_accessor_on_node(g, nid, true, 0, std::move(accessor));
+ return nid;
+}
+
+NodeID GraphBuilder::add_output_node(Graph &g, NodeParams params, NodeIdxPair input, ITensorAccessorUPtr accessor)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<OutputNode>();
+ g.add_connection(input.node_id, input.index, nid, 0);
+ set_node_params(g, nid, params);
+ set_accessor_on_node(g, nid, false, 0, std::move(accessor));
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_activation_node(Graph &g, NodeParams params, NodeIdxPair input, ActivationLayerInfo act_info)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<ActivationLayerNode>(act_info);
+ g.add_connection(input.node_id, input.index, nid, 0);
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_batch_normalization_node(Graph &g, NodeParams params, NodeIdxPair input, float epsilon,
+ ITensorAccessorUPtr mean_accessor, ITensorAccessorUPtr var_accessor,
+ ITensorAccessorUPtr beta_accessor, ITensorAccessorUPtr gamma_accessor)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ bool has_beta = (beta_accessor != nullptr);
+ bool has_gamma = (gamma_accessor != nullptr);
+
+ // Get input tensor descriptor
+ const TensorDescriptor input_tensor_desc = get_tensor_descriptor(g, g.node(input.node_id)->outputs()[0]);
+
+ // Calculate Common Descriptor
+ TensorDescriptor common_desc = input_tensor_desc;
+ common_desc.shape = TensorShape(common_desc.shape.z());
+
+ // Create mean and nodes
+ auto mean_nid = add_const_node_with_name(g, params, "Mean", common_desc, std::move(mean_accessor));
+ auto var_nid = add_const_node_with_name(g, params, "Variance", common_desc, std::move(var_accessor));
+
+ // Create beta node
+ NodeID beta_nid = EmptyNodeID;
+ if(has_beta)
+ {
+ beta_nid = add_const_node_with_name(g, params, "Beta", common_desc, std::move(beta_accessor));
+ }
+
+ // Create gamma node
+ NodeID gamma_nid = EmptyNodeID;
+ if(has_gamma)
+ {
+ gamma_nid = add_const_node_with_name(g, params, "Gamma", common_desc, std::move(gamma_accessor));
+ }
+
+ // Create batch normalization node and add connections
+ NodeID batch_norm_nid = g.add_node<BatchNormalizationLayerNode>(epsilon);
+ g.add_connection(input.node_id, input.index, batch_norm_nid, 0);
+ g.add_connection(mean_nid, 0, batch_norm_nid, 1);
+ g.add_connection(var_nid, 0, batch_norm_nid, 2);
+ if(has_beta)
+ {
+ g.add_connection(beta_nid, 0, batch_norm_nid, 3);
+ }
+ if(has_gamma)
+ {
+ g.add_connection(gamma_nid, 0, batch_norm_nid, 4);
+ }
+ set_node_params(g, batch_norm_nid, params);
+
+ return batch_norm_nid;
+}
+
+NodeID GraphBuilder::add_convolution_node(Graph &g, NodeParams params, NodeIdxPair input,
+ Size2D kernel_spatial_extend, unsigned int depth, PadStrideInfo conv_info,
+ ConvolutionMethod method,
+ ITensorAccessorUPtr weights_accessor, ITensorAccessorUPtr bias_accessor)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+ ARM_COMPUTE_ERROR_ON(depth == 0);
+ ARM_COMPUTE_ERROR_ON((kernel_spatial_extend.width == 0) || (kernel_spatial_extend.height == 0));
+
+ bool has_bias = (bias_accessor != nullptr);
+
+ // Get input tensor descriptor
+ const TensorDescriptor input_tensor_desc = get_tensor_descriptor(g, g.node(input.node_id)->outputs()[0]);
+
+ // Create weights node
+ TensorDescriptor w_desc = input_tensor_desc;
+ w_desc.shape = TensorShape(kernel_spatial_extend.width, kernel_spatial_extend.height, w_desc.shape.z(), depth);
+ NodeID w_nid = add_const_node_with_name(g, params, "Weights", w_desc, std::move(weights_accessor));
+
+ // Create bias nodes
+ NodeID b_nid = EmptyNodeID;
+ if(has_bias)
+ {
+ TensorDescriptor b_desc = input_tensor_desc;
+ b_desc.shape = TensorShape(depth);
+ b_nid = add_const_node_with_name(g, params, "Bias", b_desc, std::move(bias_accessor));
+ }
+
+ // Create convolution node and connect
+ NodeID conv_nid = g.add_node<ConvolutionLayerNode>(conv_info, method);
+ g.add_connection(input.node_id, input.index, conv_nid, 0);
+ g.add_connection(w_nid, 0, conv_nid, 1);
+ if(has_bias)
+ {
+ g.add_connection(b_nid, 0, conv_nid, 2);
+ }
+ set_node_params(g, conv_nid, params);
+
+ return conv_nid;
+}
+
+NodeID GraphBuilder::add_depth_concatenate_node(Graph &g, NodeParams params, std::vector<NodeIdxPair> inputs)
+{
+ ARM_COMPUTE_ERROR_ON(inputs.size() == 0);
+
+ NodeID nid = g.add_node<DepthConcatenateLayerNode>(inputs.size());
+
+ unsigned int i = 0;
+ for(const auto &input : inputs)
+ {
+ CHECK_NODEIDX_PAIR(input, g);
+ g.add_connection(input.node_id, input.index, nid, i++);
+ }
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_depthwise_convolution_node(Graph &g, NodeParams params, NodeIdxPair input, Size2D kernel_spatial_extend, PadStrideInfo conv_info,
+ DepthwiseConvolutionMethod method,
+ ITensorAccessorUPtr weights_accessor, ITensorAccessorUPtr bias_accessor)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+ ARM_COMPUTE_ERROR_ON((kernel_spatial_extend.width == 0) || (kernel_spatial_extend.height == 0));
+
+ bool has_bias = (bias_accessor != nullptr);
+
+ // Get input tensor descriptor
+ const TensorDescriptor input_tensor_desc = get_tensor_descriptor(g, g.node(input.node_id)->outputs()[0]);
+
+ // Create weights node
+ TensorDescriptor w_desc = input_tensor_desc;
+ w_desc.shape = TensorShape(kernel_spatial_extend.width, kernel_spatial_extend.height, w_desc.shape.z());
+ NodeID w_nid = add_const_node_with_name(g, params, "Weights", w_desc, std::move(weights_accessor));
+
+ // Create bias nodes
+ NodeID b_nid = EmptyNodeID;
+ if(has_bias)
+ {
+ TensorDescriptor b_desc = input_tensor_desc;
+ b_desc.shape = TensorShape(b_desc.shape.z());
+ b_nid = add_const_node_with_name(g, params, "Bias", b_desc, std::move(bias_accessor));
+ }
+
+ // Create convolution node and connect
+ NodeID conv_nid = g.add_node<DepthwiseConvolutionLayerNode>(conv_info, method);
+ g.add_connection(input.node_id, input.index, conv_nid, 0);
+ g.add_connection(w_nid, 0, conv_nid, 1);
+ if(has_bias)
+ {
+ g.add_connection(b_nid, 0, conv_nid, 2);
+ }
+ set_node_params(g, conv_nid, params);
+
+ return conv_nid;
+}
+
+NodeID GraphBuilder::add_elementwise_node(Graph &g, NodeParams params, NodeIdxPair input0, NodeIdxPair input1, EltwiseOperation operation)
+{
+ CHECK_NODEIDX_PAIR(input0, g);
+ CHECK_NODEIDX_PAIR(input1, g);
+
+ NodeID nid = g.add_node<EltwiseLayerNode>(operation);
+
+ g.add_connection(input0.node_id, input0.index, nid, 0);
+ g.add_connection(input1.node_id, input1.index, nid, 1);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_flatten_node(Graph &g, NodeParams params, NodeIdxPair input)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<FlattenLayerNode>();
+ g.add_connection(input.node_id, input.index, nid, 0);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_fully_connected_layer(Graph &g, NodeParams params, NodeIdxPair input, unsigned int num_outputs,
+ ITensorAccessorUPtr weights_accessor, ITensorAccessorUPtr bias_accessor)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+ ARM_COMPUTE_ERROR_ON(num_outputs == 0);
+
+ bool has_bias = (bias_accessor != nullptr);
+
+ // Get input tensor descriptor
+ const TensorDescriptor input_tensor_desc = get_tensor_descriptor(g, g.node(input.node_id)->outputs()[0]);
+
+ // Create weights node
+ TensorDescriptor w_desc = input_tensor_desc;
+ w_desc.shape = FullyConnectedLayerNode::compute_weights_shape(input_tensor_desc.shape, num_outputs);
+ NodeID w_nid = add_const_node_with_name(g, params, "Weights", w_desc, std::move(weights_accessor));
+
+ // Create bias nodes
+ NodeID b_nid = EmptyNodeID;
+ if(has_bias)
+ {
+ TensorDescriptor b_desc = input_tensor_desc;
+ b_desc.shape = TensorShape(num_outputs);
+ b_nid = add_const_node_with_name(g, params, "Bias", b_desc, std::move(bias_accessor));
+ }
+
+ // Create convolution node and connect
+ NodeID fc_nid = g.add_node<FullyConnectedLayerNode>(num_outputs);
+ g.add_connection(input.node_id, input.index, fc_nid, 0);
+ g.add_connection(w_nid, 0, fc_nid, 1);
+ if(has_bias)
+ {
+ g.add_connection(b_nid, 0, fc_nid, 2);
+ }
+
+ set_node_params(g, fc_nid, params);
+
+ return fc_nid;
+}
+
+NodeID GraphBuilder::add_normalization_node(Graph &g, NodeParams params, NodeIdxPair input, NormalizationLayerInfo norm_info)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<NormalizationLayerNode>(norm_info);
+ g.add_connection(input.node_id, input.index, nid, 0);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_pooling_node(Graph &g, NodeParams params, NodeIdxPair input, PoolingLayerInfo pool_info)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<PoolingLayerNode>(pool_info);
+ g.add_connection(input.node_id, input.index, nid, 0);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_reshape_node(Graph &g, NodeParams params, NodeIdxPair input, TensorShape shape)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<ReshapeLayerNode>(shape);
+ g.add_connection(input.node_id, input.index, nid, 0);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+
+NodeID GraphBuilder::add_softmax_node(Graph &g, NodeParams params, NodeIdxPair input, float beta)
+{
+ CHECK_NODEIDX_PAIR(input, g);
+
+ NodeID nid = g.add_node<SoftmaxLayerNode>(beta);
+ g.add_connection(input.node_id, input.index, nid, 0);
+
+ set_node_params(g, nid, params);
+
+ return nid;
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/GraphContext.cpp b/src/graph2/GraphContext.cpp
new file mode 100644
index 0000000000..88fc5216a4
--- /dev/null
+++ b/src/graph2/GraphContext.cpp
@@ -0,0 +1,84 @@
+/*
+ * 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 "arm_compute/graph2/GraphContext.h"
+#include <arm_compute/graph2.h>
+
+namespace arm_compute
+{
+namespace graph2
+{
+GraphContext::GraphContext()
+ : _tunable(false), _memory_managed(false), _memory_managers()
+{
+}
+
+void GraphContext::enable_tuning(bool enable_tuning)
+{
+ _tunable = enable_tuning;
+}
+
+bool GraphContext::is_tuning_enabled() const
+{
+ return _tunable;
+}
+
+void GraphContext::enable_memory_managenent(bool enable_mm)
+{
+ _memory_managed = enable_mm;
+}
+
+bool GraphContext::is_memory_management_enabled()
+{
+ return _memory_managed;
+}
+
+bool GraphContext::insert_memory_management_ctx(MemoryManagerContext &&memory_ctx)
+{
+ Target target = memory_ctx.target;
+ if(target == Target::UNSPECIFIED || _memory_managers.find(target) != std::end(_memory_managers))
+ {
+ return false;
+ }
+
+ _memory_managers[target] = std::move(memory_ctx);
+ return true;
+}
+
+MemoryManagerContext *GraphContext::memory_management_ctx(Target target)
+{
+ return (_memory_managers.find(target) != std::end(_memory_managers)) ? &_memory_managers[target] : nullptr;
+}
+
+void GraphContext::finalize()
+{
+ for(auto &mm_obj : _memory_managers)
+ {
+ if(mm_obj.second.mm != nullptr)
+ {
+ mm_obj.second.mm->finalize();
+ }
+ }
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/GraphManager.cpp b/src/graph2/GraphManager.cpp
new file mode 100644
index 0000000000..edbe2cc381
--- /dev/null
+++ b/src/graph2/GraphManager.cpp
@@ -0,0 +1,108 @@
+/*
+ * 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 "arm_compute/graph2/GraphManager.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/PassManager.h"
+#include "arm_compute/graph2/Utils.h"
+#include "arm_compute/graph2/detail/ExecutionHelpers.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+GraphManager::GraphManager()
+ : _workloads()
+{
+ detail::default_initialize_backends();
+}
+
+void GraphManager::finalize_graph(Graph &graph, GraphContext &ctx, PassManager &pm, Target target)
+{
+ // Setup graph context if not done manually
+ setup_default_graph_context(ctx);
+
+ // Check if graph has been registered
+ ARM_COMPUTE_ERROR_ON_MSG(_workloads.find(graph.id()) != std::end(_workloads), "Graph is already registered!");
+
+ // Force target to all graph construct
+ // TODO (geopin01) : Support heterogeneous execution
+ Target forced_target = is_target_supported(target) ? target : get_default_target();
+ force_target_to_graph(graph, forced_target);
+
+ // Configure all tensors
+ detail::configure_all_tensors(graph);
+
+ // Apply all mutating passes
+ pm.run_all(graph);
+
+ // TODO (geopin01): Perform a graph validation
+
+ // Perform topological sort
+ // FIXME : Sort nodes and pass sorted indices in configure all nodes
+
+ // Configure all nodes
+ auto workload = detail::configure_all_nodes(graph, ctx);
+ ARM_COMPUTE_ERROR_ON_MSG(workload.tasks.empty(), "Could not configure all nodes!");
+
+ // Allocate all tensors
+ detail::allocate_all_tensors(graph);
+
+ // Call accessors on all Const nodes
+ detail::call_all_const_node_accessors(graph);
+
+ _workloads.insert(std::make_pair(graph.id(), std::move(workload)));
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Created workload for graph with ID : " << graph.id().get() << std::endl);
+
+ // Finalize Graph context
+ ctx.finalize();
+}
+
+void GraphManager::execute_graph(Graph &graph)
+{
+ // Check if graph is finalized
+ auto it = _workloads.find(graph.id());
+ ARM_COMPUTE_ERROR_ON_MSG(it == std::end(_workloads), "Graph is not registered!");
+
+ // Call input accessors
+ detail::call_all_input_node_accessors(it->second);
+
+ // Run graph
+ detail::call_all_tasks(it->second);
+
+ // Call output accessors
+ detail::call_all_output_node_accessors(it->second);
+}
+
+void GraphManager::invalidate_graph(Graph &graph)
+{
+ auto it = _workloads.find(graph.id());
+ ARM_COMPUTE_ERROR_ON_MSG(it == std::end(_workloads), "Graph is not registered!");
+
+ _workloads.erase(it);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/INode.cpp b/src/graph2/INode.cpp
new file mode 100644
index 0000000000..28be341396
--- /dev/null
+++ b/src/graph2/INode.cpp
@@ -0,0 +1,193 @@
+/*
+ * 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 "arm_compute/graph2/INode.h"
+
+#include "arm_compute/core/Error.h"
+#include "arm_compute/graph2/Edge.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Tensor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+// *INDENT-OFF*
+// clang-format off
+INode::INode()
+ : _graph(nullptr), _id(EmptyNodeID), _common_params({ "", Target::UNSPECIFIED}),
+ _outputs(), _input_edges(), _output_edges(), _assigned_target(Target::UNSPECIFIED)
+{
+}
+// clang-format on
+// *INDENT-ON*
+
+void INode::set_graph(Graph *g)
+{
+ ARM_COMPUTE_ERROR_ON(g == nullptr);
+ _graph = g;
+}
+
+void INode::set_id(NodeID id)
+{
+ _id = id;
+}
+
+void INode::set_common_node_parameters(NodeParams common_params)
+{
+ _common_params = std::move(common_params);
+}
+
+void INode::set_requested_target(Target target)
+{
+ _common_params.target = target;
+}
+
+void INode::set_assigned_target(Target target)
+{
+ _assigned_target = target;
+}
+
+void INode::set_output_tensor(TensorID tid, size_t idx)
+{
+ if(tid != NullTensorID && (idx < _outputs.size()) && (_graph->tensor(tid) != nullptr))
+ {
+ ARM_COMPUTE_ERROR_ON(_graph == nullptr);
+ Tensor *updated_tensor = _graph->tensor(tid);
+ _outputs[idx] = tid;
+
+ // Set tensor to all output edges of the node
+ for(auto &output_edge_id : _output_edges)
+ {
+ auto output_edge = _graph->edge(output_edge_id);
+ if(output_edge != nullptr)
+ {
+ // Unbind edge from current tensor
+ auto current_output_tensor = output_edge->tensor();
+ current_output_tensor->unbind_edge(output_edge->id());
+
+ // Update tensor to edge and rebind tensor
+ output_edge->update_bound_tensor(updated_tensor);
+ updated_tensor->bind_edge(output_edge->id());
+ }
+ }
+ }
+}
+
+NodeID INode::id() const
+{
+ return _id;
+}
+
+std::string INode::name() const
+{
+ return _common_params.name;
+}
+
+const Graph *INode::graph() const
+{
+ return _graph;
+}
+
+Graph *INode::graph()
+{
+ return _graph;
+}
+
+const std::vector<TensorID> &INode::outputs() const
+{
+ return _outputs;
+}
+
+const std::vector<EdgeID> &INode::input_edges() const
+{
+ return _input_edges;
+}
+
+const std::set<EdgeID> &INode::output_edges() const
+{
+ return _output_edges;
+}
+
+TensorID INode::input_id(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(idx >= _input_edges.size());
+ Edge *e = _graph->edge(_input_edges[idx]);
+ return (e != nullptr) ? e->tensor_id() : NullTensorID;
+}
+
+TensorID INode::output_id(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+ return _outputs[idx];
+}
+
+Tensor *INode::input(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(_graph == nullptr);
+ ARM_COMPUTE_ERROR_ON(idx >= _input_edges.size());
+ Edge *e = _graph->edge(_input_edges[idx]);
+ return (e != nullptr) ? e->tensor() : nullptr;
+}
+
+Tensor *INode::output(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(_graph == nullptr);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+ return _graph->tensor(_outputs[idx]);
+}
+
+EdgeID INode::input_edge_id(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(idx >= _input_edges.size());
+ return _input_edges[idx];
+}
+
+Edge *INode::input_edge(size_t idx) const
+{
+ ARM_COMPUTE_ERROR_ON(_graph == nullptr);
+ ARM_COMPUTE_ERROR_ON(idx >= _input_edges.size());
+ return _graph->edge(_input_edges[idx]);
+}
+
+size_t INode::num_inputs() const
+{
+ return _input_edges.size();
+}
+
+size_t INode::num_outputs() const
+{
+ return _outputs.size();
+}
+
+Target INode::requested_target() const
+{
+ return _common_params.target;
+}
+
+Target INode::assigned_target() const
+{
+ return _assigned_target;
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/PassManager.cpp b/src/graph2/PassManager.cpp
new file mode 100644
index 0000000000..2fa937bd89
--- /dev/null
+++ b/src/graph2/PassManager.cpp
@@ -0,0 +1,88 @@
+/*
+ * 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 "arm_compute/graph2/PassManager.h"
+
+#include "arm_compute/graph2/Logger.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+PassManager::PassManager()
+ : _passes()
+{
+}
+
+const std::vector<std::unique_ptr<IGraphMutator>> &PassManager::passes() const
+{
+ return _passes;
+}
+
+IGraphMutator *PassManager::pass(size_t index)
+{
+ return (index >= _passes.size()) ? nullptr : _passes.at(index).get();
+}
+
+void PassManager::append(std::unique_ptr<IGraphMutator> pass)
+{
+ if(pass)
+ {
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Appending mutating pass : " << pass->name() << std::endl);
+ _passes.push_back(std::move(pass));
+ }
+}
+
+void PassManager::clear()
+{
+ _passes.clear();
+}
+
+void PassManager::run_all(Graph &g)
+{
+ for(auto &pass : _passes)
+ {
+ if(pass)
+ {
+ ARM_COMPUTE_LOG_GRAPH_INFO("Running mutating pass : " << pass->name() << std::endl);
+ pass->mutate(g);
+ }
+ }
+}
+
+void PassManager::run(Graph &g, size_t index)
+{
+ if(index >= _passes.size())
+ {
+ return;
+ }
+
+ auto &pass = _passes.at(index);
+
+ if(pass != nullptr)
+ {
+ pass->mutate(g);
+ }
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/Tensor.cpp b/src/graph2/Tensor.cpp
new file mode 100644
index 0000000000..c6054d716d
--- /dev/null
+++ b/src/graph2/Tensor.cpp
@@ -0,0 +1,111 @@
+/*
+ * 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 "arm_compute/graph2/Tensor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+Tensor::Tensor(TensorID id, TensorDescriptor desc)
+ : _id(id), _desc(desc), _handle(nullptr), _accessor(nullptr), _bound_edges()
+{
+}
+
+TensorID Tensor::id() const
+{
+ return _id;
+}
+
+TensorDescriptor &Tensor::desc()
+{
+ return _desc;
+}
+
+const TensorDescriptor &Tensor::desc() const
+{
+ return _desc;
+}
+
+void Tensor::set_handle(std::unique_ptr<ITensorHandle> backend_tensor)
+{
+ _handle = std::move(backend_tensor);
+}
+
+ITensorHandle *Tensor::handle()
+{
+ return _handle.get();
+}
+
+void Tensor::set_accessor(std::unique_ptr<ITensorAccessor> accessor)
+{
+ _accessor = std::move(accessor);
+}
+
+ITensorAccessor *Tensor::accessor()
+{
+ return _accessor.get();
+}
+
+bool Tensor::call_accessor()
+{
+ // Early exit guard
+ if(!_accessor || !_handle)
+ {
+ return false;
+ }
+
+ // Map tensor
+ _handle->map(true);
+
+ // Return in case of null backend buffer
+ if(_handle->tensor().buffer() == nullptr)
+ {
+ return false;
+ }
+
+ // Call accessor
+ _accessor->access_tensor(_handle->tensor());
+
+ // Unmap tensor
+ _handle->unmap();
+
+ return true;
+}
+
+void Tensor::bind_edge(EdgeID eid)
+{
+ _bound_edges.insert(eid);
+}
+
+void Tensor::unbind_edge(EdgeID eid)
+{
+ _bound_edges.erase(eid);
+}
+
+const std::set<EdgeID> Tensor::bound_edges() const
+{
+ return _bound_edges;
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/Utils.cpp b/src/graph2/Utils.cpp
new file mode 100644
index 0000000000..a518c80da8
--- /dev/null
+++ b/src/graph2/Utils.cpp
@@ -0,0 +1,97 @@
+/*
+ * 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 "arm_compute/graph2/Utils.h"
+
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/backends/BackendRegistry.h"
+#include "arm_compute/graph2/mutators/GraphMutators.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+bool is_target_supported(Target target)
+{
+ return backends::BackendRegistry::get().contains(target);
+}
+
+Target get_default_target()
+{
+ if(is_target_supported(Target::NEON))
+ {
+ return Target::NEON;
+ }
+ if(is_target_supported(Target::CL))
+ {
+ return Target::CL;
+ }
+
+ ARM_COMPUTE_ERROR("No backend exists!");
+}
+
+void force_target_to_graph(Graph &g, Target target)
+{
+ auto &nodes = g.nodes();
+ for(auto &node : nodes)
+ {
+ if(node)
+ {
+ node->set_assigned_target(target);
+ }
+ }
+
+ auto &tensors = g.tensors();
+ for(auto &tensor : tensors)
+ {
+ if(tensor)
+ {
+ tensor->desc().target = target;
+ }
+ }
+}
+
+PassManager create_default_pass_manager()
+{
+ PassManager pm;
+
+ pm.append(support::cpp14::make_unique<InPlaceOperationMutator>());
+ pm.append(support::cpp14::make_unique<NodeFusionMutator>());
+ pm.append(support::cpp14::make_unique<DepthConcatSubTensorMutator>());
+
+ return pm;
+}
+
+/** Default setups a graph Context
+ *
+ * @param[in] ctx Context to default initialize
+ */
+void setup_default_graph_context(GraphContext &ctx)
+{
+ for(const auto &backend : backends::BackendRegistry::get().backends())
+ {
+ backend.second->setup_backend_context(ctx);
+ }
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/Workload.cpp b/src/graph2/Workload.cpp
new file mode 100644
index 0000000000..3fd36fabc7
--- /dev/null
+++ b/src/graph2/Workload.cpp
@@ -0,0 +1,41 @@
+/*
+ * 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 "arm_compute/graph2/Workload.h"
+
+#include "arm_compute/graph2/INode.h"
+#include "arm_compute/graph2/ITensorHandle.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+void ExecutionTask::operator()()
+{
+ if(task)
+ {
+ task->run();
+ }
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/BackendRegistry.cpp b/src/graph2/backends/BackendRegistry.cpp
new file mode 100644
index 0000000000..5f1218f335
--- /dev/null
+++ b/src/graph2/backends/BackendRegistry.cpp
@@ -0,0 +1,63 @@
+/*
+ * 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 "arm_compute/graph2/backends/BackendRegistry.h"
+
+using namespace arm_compute::graph2::backends;
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+BackendRegistry::BackendRegistry()
+ : _registered_backends()
+{
+}
+
+BackendRegistry &BackendRegistry::get()
+{
+ static BackendRegistry instance;
+ return instance;
+}
+
+IDeviceBackend *BackendRegistry::find_backend(Target target)
+{
+ ARM_COMPUTE_ERROR_ON(!contains(target));
+ return _registered_backends[target].get();
+}
+
+bool BackendRegistry::contains(Target target) const
+{
+ auto it = _registered_backends.find(target);
+ return (it != _registered_backends.end());
+}
+
+const std::map<Target, std::unique_ptr<IDeviceBackend>> &BackendRegistry::backends() const
+{
+ return _registered_backends;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/backends/CL/CLDeviceBackend.cpp b/src/graph2/backends/CL/CLDeviceBackend.cpp
new file mode 100644
index 0000000000..e06033121a
--- /dev/null
+++ b/src/graph2/backends/CL/CLDeviceBackend.cpp
@@ -0,0 +1,175 @@
+/*
+ * 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 "arm_compute/graph2/backends/CL/CLDeviceBackend.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/INode.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/Tensor.h"
+#include "arm_compute/graph2/backends/BackendRegistrar.h"
+#include "arm_compute/graph2/backends/CL/CLFunctionFactory.h"
+#include "arm_compute/graph2/backends/CL/CLSubTensorHandle.h"
+#include "arm_compute/graph2/backends/CL/CLTensorHandle.h"
+
+#include "arm_compute/core/TensorInfo.h"
+#include "arm_compute/runtime/BlobLifetimeManager.h"
+#include "arm_compute/runtime/CL/CLBufferAllocator.h"
+#include "arm_compute/runtime/CL/CLScheduler.h"
+#include "arm_compute/runtime/MemoryManagerOnDemand.h"
+#include "arm_compute/runtime/PoolManager.h"
+
+#include "support/ToolchainSupport.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+namespace
+{
+bool file_exists(const std::string &filename)
+{
+ std::ifstream file(filename);
+ return file.good();
+}
+} // namespace
+
+/** Register CL backend */
+static detail::BackendRegistrar<CLDeviceBackend> CLDeviceBackend_registrar(Target::CL);
+
+/** Tuner export file */
+static const std::string tuner_data_filename = "acl_tuner.csv";
+
+CLDeviceBackend::CLDeviceBackend()
+ : _tuner(), _allocator(cl::Context::getDefault())
+{
+}
+
+CLDeviceBackend::~CLDeviceBackend()
+{
+ // TODO (geopin01) : Shouldn't call non exception safe stuff here
+ if(_tuner.tune_new_kernels() && !_tuner.lws_table().empty())
+ {
+ _tuner.save_to_file(tuner_data_filename);
+ }
+}
+
+void CLDeviceBackend::set_kernel_tuning(bool enable_tuning)
+{
+ _tuner.set_tune_new_kernels(enable_tuning);
+}
+
+void CLDeviceBackend::initialize_backend()
+{
+ // Load tuner data if available
+ if(_tuner.lws_table().empty() && file_exists(tuner_data_filename))
+ {
+ _tuner.load_from_file(tuner_data_filename);
+ }
+
+ // Setup Scheduler
+ CLScheduler::get().default_init(&_tuner);
+
+ // Create allocator with new context
+ _allocator = CLBufferAllocator();
+}
+
+void CLDeviceBackend::setup_backend_context(GraphContext &ctx)
+{
+ // Setup tuner
+ set_kernel_tuning(ctx.is_tuning_enabled());
+
+ // Setup a management backend
+ if(ctx.memory_management_ctx(Target::CL) == nullptr)
+ {
+ MemoryManagerContext mm_ctx;
+ mm_ctx.target = Target::CL;
+ mm_ctx.mm = create_memory_manager(MemoryManagerAffinity::Buffer);
+
+ ctx.insert_memory_management_ctx(std::move(mm_ctx));
+ }
+}
+
+std::unique_ptr<ITensorHandle> CLDeviceBackend::create_tensor(const Tensor &tensor)
+{
+ // Get tensor descriptor
+ const TensorDescriptor &tensor_desc = tensor.desc();
+ ARM_COMPUTE_ERROR_ON(tensor_desc.target != Target::CL);
+
+ // Create backend tensor handle
+ TensorInfo info(tensor_desc.shape, 1, tensor_desc.data_type);
+ auto backend_tensor_handle = support::cpp14::make_unique<CLTensorHandle>(info);
+
+ return std::move(backend_tensor_handle);
+}
+
+std::unique_ptr<ITensorHandle> CLDeviceBackend::create_subtensor(ITensorHandle *parent, TensorShape shape, Coordinates coords)
+{
+ if(parent == nullptr)
+ {
+ return nullptr;
+ }
+
+ return support::cpp14::make_unique<CLSubTensorHandle>(parent, shape, coords);
+}
+
+std::unique_ptr<arm_compute::IFunction> CLDeviceBackend::configure_node(INode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Configuring CL node with ID : " << node.id() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.assigned_target() != Target::CL);
+
+ // Configure node
+ return CLFunctionFactory::create(&node, ctx);
+}
+
+arm_compute::Status CLDeviceBackend::validate_node(const INode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Validating CL node with ID : " << node.id() << std::endl);
+
+ ARM_COMPUTE_UNUSED(node);
+
+ return Status{};
+}
+
+std::shared_ptr<arm_compute::IMemoryManager> CLDeviceBackend::create_memory_manager(MemoryManagerAffinity affinity)
+{
+ if(affinity == MemoryManagerAffinity::Offset)
+ {
+ ARM_COMPUTE_LOG_GRAPH_WARNING("CL Backend does not support offset affinity memory management!");
+ return nullptr;
+ }
+
+ auto lifetime_mgr = std::make_shared<BlobLifetimeManager>();
+ auto pool_mgr = std::make_shared<PoolManager>();
+ auto mm = std::make_shared<MemoryManagerOnDemand>(lifetime_mgr, pool_mgr);
+
+ mm->set_allocator(&_allocator);
+
+ return mm;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/CL/CLFunctionsFactory.cpp b/src/graph2/backends/CL/CLFunctionsFactory.cpp
new file mode 100644
index 0000000000..bba0cce025
--- /dev/null
+++ b/src/graph2/backends/CL/CLFunctionsFactory.cpp
@@ -0,0 +1,584 @@
+/*
+ * 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 "arm_compute/graph2/backends/CL/CLFunctionFactory.h"
+
+#include "arm_compute/core/utils/misc/Cast.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/TypePrinter.h"
+#include "arm_compute/graph2/Types.h"
+#include "arm_compute/graph2/backends/Utils.h"
+#include "arm_compute/graph2/nodes/Nodes.h"
+#include "arm_compute/runtime/CL/CLFunctions.h"
+
+#include "support/ToolchainSupport.h"
+
+using namespace arm_compute::utils::cast;
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+namespace
+{
+/** Returns backing tensor of a given tensor
+ *
+ * @param[in] tensor Tensor to extract the backing tensor from
+ *
+ * @return Backing tensor if present else nullptr
+ */
+arm_compute::ICLTensor *get_backing_tensor(arm_compute::graph2::Tensor *tensor)
+{
+ arm_compute::ICLTensor *backing_tensor = nullptr;
+ if(tensor != nullptr)
+ {
+ ARM_COMPUTE_ERROR_ON(tensor->desc().target != arm_compute::graph2::Target::CL);
+ // Get backing tensor handle
+ ITensorHandle *tensor_handle = tensor->handle();
+ // Get backing tensor
+ backing_tensor = (tensor_handle != nullptr) ? polymorphic_cast<ICLTensor *>(&tensor_handle->tensor()) : nullptr;
+ }
+
+ return backing_tensor;
+}
+
+/** Create a backend activation layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend activation layer function
+ */
+std::unique_ptr<IFunction> create_activation_layer(ActivationLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL ActivationLayerNode node with ID : " << node.id() << " and Name: " << node.name()
+ << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const ActivationLayerInfo act_info = node.activation_info();
+
+ // Create function
+ auto func = support::cpp14::make_unique<CLActivationLayer>();
+ func->configure(input, output, act_info);
+
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLActivationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Shape: " << input->info()->tensor_shape()
+ << " Activation function: " << act_info.activation()
+ << " a: " << act_info.a()
+ << " b: " << act_info.b()
+ << " InPlace : " << is_in_place_operation(input, output)
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend batch normalization layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend batch normalization layer function
+ */
+std::unique_ptr<IFunction> create_batch_normalization_layer(BatchNormalizationLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating CL BatchNormalization node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+
+ // TODO (geopin01) : Var and mean are compulsory, switch function to accept nullptr as beta and/or gamma
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 5);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *mean = get_backing_tensor(node.input(1));
+ ICLTensor *var = get_backing_tensor(node.input(2));
+ ICLTensor *beta = get_backing_tensor(node.input(3));
+ ICLTensor *gamma = get_backing_tensor(node.input(4));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const float epsilon = node.epsilon();
+ const ActivationLayerInfo fused_act = node.fused_activation();
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLBatchNormalizationLayer>();
+ func->configure(input, output, mean, var, beta, gamma, epsilon, fused_act);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLBatchNormalizationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Shape: " << input->info()->tensor_shape()
+ << " Epsilon: " << epsilon << " "
+ << (fused_act.enabled() ? to_string(fused_act.activation()) : "")
+ << " InPlace : " << is_in_place_operation(input, output)
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend convolution layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend convolution layer function
+ */
+std::unique_ptr<IFunction> create_convolution_layer(ConvolutionLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating CL ConvolutionLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *weights = get_backing_tensor(node.input(1));
+ ICLTensor *biases = get_backing_tensor(node.input(2));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const PadStrideInfo conv_info = node.convolution_info();
+ const ConvolutionMethod conv_algorithm = node.convolution_method();
+
+ // Create and configure function (we assume that functions have been validated before creation)
+ std::shared_ptr<IMemoryManager> mm = get_memory_manager(ctx, Target::CL);
+ std::unique_ptr<IFunction> func;
+ std::string func_name;
+ if(conv_algorithm == ConvolutionMethod::DIRECT)
+ {
+ std::tie(func, func_name) = create_named_function<CLDirectConvolutionLayer>(
+ std::string("CLDirectConvolutionLayer"), input, weights, biases, output, conv_info);
+ }
+ else if(conv_algorithm == ConvolutionMethod::GEMM)
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<CLGEMMConvolutionLayer>(std::string("CLGEMMConvolutionLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+ else
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<CLConvolutionLayer>(std::string("CLConvolutionLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+ return func;
+}
+
+/** Create a backend layer depth concatenate function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend depth concatenate layer function
+ */
+std::unique_ptr<arm_compute::IFunction> create_depth_concatenate_layer(DepthConcatenateLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating CL DepthConcatenate node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Return nullptr if depth concatenate is switched off
+ if(!node.is_enabled())
+ {
+ return nullptr;
+ }
+
+ // Extract IO and info
+ std::vector<arm_compute::ICLTensor *> inputs;
+ for(unsigned int i = 0; i < node.num_inputs(); ++i)
+ {
+ inputs.push_back(get_backing_tensor(node.input(i)));
+ }
+ ICLTensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLDepthConcatenateLayer>();
+ func->configure(inputs, output);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLDepthConcatenateLayer"
+ << " Data Type: " << output->info()->data_type()
+ << " Shape: " << output->info()->tensor_shape()
+ << " Num Inputs: " << inputs.size()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend layer depth-wise convolution function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend depth-wise convolution layer function
+ */
+std::unique_ptr<IFunction> create_depthwise_convolution_layer(DepthwiseConvolutionLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL DepthwiseConvolutionLayer node with ID : " << node.id() << " and Name: " << node.name()
+ << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *weights = get_backing_tensor(node.input(1));
+ ICLTensor *biases = get_backing_tensor(node.input(2));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const PadStrideInfo conv_info = node.convolution_info();
+ const DepthwiseConvolutionMethod dwc_algorithm = node.depthwise_convolution_method();
+
+ // Create and configure function (we assume that functions have been validated before creation)
+ std::unique_ptr<IFunction> func;
+ std::string func_name;
+ if(dwc_algorithm == DepthwiseConvolutionMethod::OPTIMIZED_3x3)
+ {
+ std::tie(func, func_name) = create_named_function<CLDepthwiseConvolutionLayer3x3>(
+ std::string("CLDepthwiseConvolutionLayer3x3"), input, weights, biases, output, conv_info);
+ }
+ else
+ {
+ std::tie(func, func_name) = create_named_function<CLDepthwiseConvolutionLayer>(
+ std::string("CLDepthwiseConvolutionLayer"), input, weights, biases, output, conv_info);
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+ return func;
+}
+
+/** Create a backend element-wise operation layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend element-wise operation layer function
+ */
+std::unique_ptr<IFunction> create_eltwise_layer(EltwiseLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL EltwiseLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 2);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input1 = get_backing_tensor(node.input(0));
+ ICLTensor *input2 = get_backing_tensor(node.input(1));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const EltwiseOperation eltwise_op = node.eltwise_operation();
+ ARM_COMPUTE_ERROR_ON(input1 == nullptr);
+ ARM_COMPUTE_ERROR_ON(input2 == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ std::unique_ptr<IFunction> func = nullptr;
+ std::string func_name;
+ if(eltwise_op == EltwiseOperation::ADD)
+ {
+ std::tie(func, func_name) = create_named_function<CLArithmeticAddition>(std::string("CLArithmeticAddition"),
+ input1, input2, output,
+ ConvertPolicy::SATURATE);
+ }
+ else if(eltwise_op == EltwiseOperation::SUB)
+ {
+ std::tie(func, func_name) = create_named_function<CLArithmeticSubtraction>(
+ std::string("CLArithmeticSubtraction"), input1, input2, output, ConvertPolicy::SATURATE);
+ }
+ else if(eltwise_op == EltwiseOperation::MUL)
+ {
+ std::tie(func, func_name) = create_named_function<CLPixelWiseMultiplication>(
+ std::string("CLPixelWiseMultiplication"), input1, input2, output, 1.f, ConvertPolicy::SATURATE,
+ RoundingPolicy::TO_NEAREST_EVEN);
+ }
+ else
+ {
+ ARM_COMPUTE_ERROR("Unsupported element-wise operation!");
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input1->info()->data_type()
+ << " Shape : " << input1->info()->tensor_shape()
+ << std::endl);
+
+ return func;
+}
+
+/** Create a backend flatten layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend flatten layer function
+ */
+std::unique_ptr<IFunction> create_flatten_layer(FlattenLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL FlattenLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLFlattenLayer>();
+ func->configure(input, output);
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLFlattenLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend fully connected layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend fully connected layer function
+ */
+std::unique_ptr<IFunction> create_fully_connected_layer(FullyConnectedLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL FullyConnectedLayer node with ID : " << node.id() << " and Name: " << node.name()
+ << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *weights = get_backing_tensor(node.input(1));
+ ICLTensor *biases = get_backing_tensor(node.input(2));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLFullyConnectedLayer>(get_memory_manager(ctx, Target::CL));
+ func->configure(input, weights, biases, output);
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(weights == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLFullyConnectedLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Biases Shape: " << biases->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend normalization layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend normalization layer function
+ */
+std::unique_ptr<IFunction> create_normalization_layer(NormalizationLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL NormalizationLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const NormalizationLayerInfo norm_info = node.normalization_info();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLNormalizationLayer>();
+ func->configure(input, output, norm_info);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLNormalizationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << " Normalization info: " << norm_info.type()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend pooling layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend pooling layer function
+ */
+std::unique_ptr<IFunction> create_pooling_layer(PoolingLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL PoolingLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const PoolingLayerInfo pool_info = node.pooling_info();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLPoolingLayer>();
+ func->configure(input, output, pool_info);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLPoolingLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << " Pooling info: " << pool_info.pool_type()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend reshape layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend reshape layer function
+ */
+std::unique_ptr<IFunction> create_reshape_layer(ReshapeLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL ReshapeLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLReshapeLayer>();
+ func->configure(input, output);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLReshapeLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend softmax layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend softmax layer function
+ */
+std::unique_ptr<IFunction> create_softmax_layer(SoftmaxLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE(
+ "Creating CL SoftmaxLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ICLTensor *input = get_backing_tensor(node.input(0));
+ ICLTensor *output = get_backing_tensor(node.output(0));
+ const float beta = node.beta();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<CLSoftmaxLayer>(get_memory_manager(ctx, Target::CL));
+ func->configure(input, output, beta);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated CLSoftmaxLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+} // namespace
+
+std::unique_ptr<IFunction> CLFunctionFactory::create(INode *node, GraphContext &ctx)
+{
+ if(node == nullptr)
+ {
+ return nullptr;
+ }
+
+ NodeType type = node->type();
+ switch(type)
+ {
+ case NodeType::ActivationLayer:
+ return create_activation_layer(*polymorphic_downcast<ActivationLayerNode *>(node));
+ case NodeType::BatchNormalizationLayer:
+ return create_batch_normalization_layer(*polymorphic_downcast<BatchNormalizationLayerNode *>(node));
+ case NodeType::ConvolutionLayer:
+ return create_convolution_layer(*polymorphic_downcast<ConvolutionLayerNode *>(node), ctx);
+ case NodeType::DepthConcatenateLayer:
+ return create_depth_concatenate_layer(*polymorphic_downcast<DepthConcatenateLayerNode *>(node));
+ case NodeType::DepthwiseConvolutionLayer:
+ return create_depthwise_convolution_layer(*polymorphic_downcast<DepthwiseConvolutionLayerNode *>(node));
+ case NodeType::EltwiseLayer:
+ return create_eltwise_layer(*polymorphic_downcast<EltwiseLayerNode *>(node));
+ case NodeType::FlattenLayer:
+ return create_flatten_layer(*polymorphic_downcast<FlattenLayerNode *>(node));
+ case NodeType::FullyConnectedLayer:
+ return create_fully_connected_layer(*polymorphic_downcast<FullyConnectedLayerNode *>(node), ctx);
+ case NodeType::NormalizationLayer:
+ return create_normalization_layer(*polymorphic_downcast<NormalizationLayerNode *>(node));
+ case NodeType::PoolingLayer:
+ return create_pooling_layer(*polymorphic_downcast<PoolingLayerNode *>(node));
+ case NodeType::ReshapeLayer:
+ return create_reshape_layer(*polymorphic_downcast<ReshapeLayerNode *>(node));
+ case NodeType::SoftmaxLayer:
+ return create_softmax_layer(*polymorphic_downcast<SoftmaxLayerNode *>(node), ctx);
+ default:
+ return nullptr;
+ }
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/CL/CLSubTensorHandle.cpp b/src/graph2/backends/CL/CLSubTensorHandle.cpp
new file mode 100644
index 0000000000..2954652d71
--- /dev/null
+++ b/src/graph2/backends/CL/CLSubTensorHandle.cpp
@@ -0,0 +1,73 @@
+/*
+ * 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 "arm_compute/graph2/backends/CL/CLSubTensorHandle.h"
+
+#include "arm_compute/core/utils/misc/Cast.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+CLSubTensorHandle::CLSubTensorHandle(ITensorHandle *parent_handle, const TensorShape &shape, const Coordinates &coords)
+ : _sub_tensor()
+{
+ ARM_COMPUTE_ERROR_ON(!parent_handle);
+ auto parent_tensor = arm_compute::utils::cast::polymorphic_downcast<ICLTensor *>(&parent_handle->tensor());
+ _sub_tensor = arm_compute::CLSubTensor(parent_tensor, shape, coords);
+}
+
+void CLSubTensorHandle::allocate()
+{
+ // noop
+}
+
+const arm_compute::ITensor &CLSubTensorHandle::tensor() const
+{
+ return _sub_tensor;
+}
+
+arm_compute::ITensor &CLSubTensorHandle::tensor()
+{
+ return _sub_tensor;
+}
+
+void CLSubTensorHandle::map(bool blocking)
+{
+ _sub_tensor.map(blocking);
+}
+
+void CLSubTensorHandle::unmap()
+{
+ _sub_tensor.unmap();
+}
+
+bool CLSubTensorHandle::is_subtensor() const
+{
+ return true;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/CL/CLTensorHandle.cpp b/src/graph2/backends/CL/CLTensorHandle.cpp
new file mode 100644
index 0000000000..f515e0bac3
--- /dev/null
+++ b/src/graph2/backends/CL/CLTensorHandle.cpp
@@ -0,0 +1,69 @@
+/*
+ * 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 "arm_compute/graph2/backends/CL/CLTensorHandle.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+CLTensorHandle::CLTensorHandle(const ITensorInfo &info)
+ : _tensor()
+{
+ _tensor.allocator()->init(info);
+}
+
+void CLTensorHandle::allocate()
+{
+ _tensor.allocator()->allocate();
+}
+
+const arm_compute::ITensor &CLTensorHandle::tensor() const
+{
+ return _tensor;
+}
+
+arm_compute::ITensor &CLTensorHandle::tensor()
+{
+ return _tensor;
+}
+
+void CLTensorHandle::map(bool blocking)
+{
+ _tensor.map(blocking);
+}
+
+void CLTensorHandle::unmap()
+{
+ _tensor.unmap();
+}
+
+bool CLTensorHandle::is_subtensor() const
+{
+ return false;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/NEON/NEDeviceBackend.cpp b/src/graph2/backends/NEON/NEDeviceBackend.cpp
new file mode 100644
index 0000000000..9f24498abd
--- /dev/null
+++ b/src/graph2/backends/NEON/NEDeviceBackend.cpp
@@ -0,0 +1,135 @@
+/*
+ * 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 "arm_compute/graph2/backends/NEON/NEDeviceBackend.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/INode.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/Tensor.h"
+#include "arm_compute/graph2/backends/BackendRegistrar.h"
+#include "arm_compute/graph2/backends/NEON/NEFunctionFactory.h"
+#include "arm_compute/graph2/backends/NEON/NESubTensorHandle.h"
+#include "arm_compute/graph2/backends/NEON/NETensorHandle.h"
+
+#include "arm_compute/core/TensorInfo.h"
+#include "arm_compute/runtime/Allocator.h"
+#include "arm_compute/runtime/BlobLifetimeManager.h"
+#include "arm_compute/runtime/MemoryManagerOnDemand.h"
+#include "arm_compute/runtime/OffsetLifetimeManager.h"
+#include "arm_compute/runtime/PoolManager.h"
+
+#include "support/ToolchainSupport.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+/** Register NEON backend */
+static detail::BackendRegistrar<NEDeviceBackend> NEDeviceBackend_registrar(Target::NEON);
+
+NEDeviceBackend::NEDeviceBackend()
+ : _allocator()
+{
+}
+
+void NEDeviceBackend::initialize_backend()
+{
+}
+
+void NEDeviceBackend::setup_backend_context(GraphContext &ctx)
+{
+ if(ctx.memory_management_ctx(Target::NEON) == nullptr)
+ {
+ MemoryManagerContext mm_ctx;
+ mm_ctx.target = Target::NEON;
+ mm_ctx.mm = create_memory_manager(MemoryManagerAffinity::Buffer);
+
+ ctx.insert_memory_management_ctx(std::move(mm_ctx));
+ }
+}
+
+std::unique_ptr<ITensorHandle> NEDeviceBackend::create_tensor(const Tensor &tensor)
+{
+ // Get tensor descriptor
+ const TensorDescriptor &tensor_desc = tensor.desc();
+ ARM_COMPUTE_ERROR_ON(tensor_desc.target != Target::NEON);
+
+ // Create backend tensor handle
+ TensorInfo info(tensor_desc.shape, 1, tensor_desc.data_type);
+ auto backend_tensor_handle = support::cpp14::make_unique<NETensorHandle>(info);
+
+ return std::move(backend_tensor_handle);
+}
+
+std::unique_ptr<ITensorHandle> NEDeviceBackend::create_subtensor(ITensorHandle *parent, TensorShape shape, Coordinates coords)
+{
+ if(parent == nullptr)
+ {
+ return nullptr;
+ }
+
+ return support::cpp14::make_unique<NESubTensorHandle>(parent, shape, coords);
+}
+
+std::unique_ptr<arm_compute::IFunction> NEDeviceBackend::configure_node(INode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Configuring NEON node with ID : " << node.id() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.assigned_target() != Target::NEON);
+
+ // Configure node
+ return NEFunctionFactory::create(&node, ctx);
+}
+
+arm_compute::Status NEDeviceBackend::validate_node(const INode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Validating NEON node with ID : " << node.id() << std::endl);
+ ARM_COMPUTE_UNUSED(node);
+
+ return Status{};
+}
+
+std::shared_ptr<arm_compute::IMemoryManager> NEDeviceBackend::create_memory_manager(MemoryManagerAffinity affinity)
+{
+ std::shared_ptr<ILifetimeManager> lifetime_mgr = nullptr;
+ if(affinity == MemoryManagerAffinity::Buffer)
+ {
+ lifetime_mgr = std::make_shared<BlobLifetimeManager>();
+ }
+ else
+ {
+ lifetime_mgr = std::make_shared<OffsetLifetimeManager>();
+ }
+ auto pool_mgr = std::make_shared<PoolManager>();
+ auto mm = std::make_shared<MemoryManagerOnDemand>(lifetime_mgr, pool_mgr);
+
+ mm->set_allocator(&_allocator);
+
+ return mm;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/NEON/NEFunctionFactory.cpp b/src/graph2/backends/NEON/NEFunctionFactory.cpp
new file mode 100644
index 0000000000..933210377d
--- /dev/null
+++ b/src/graph2/backends/NEON/NEFunctionFactory.cpp
@@ -0,0 +1,563 @@
+/*
+ * 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 "arm_compute/graph2/backends/NEON/NEFunctionFactory.h"
+
+#include "arm_compute/core/utils/misc/Cast.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/TypePrinter.h"
+#include "arm_compute/graph2/backends/Utils.h"
+#include "arm_compute/graph2/nodes/Nodes.h"
+#include "arm_compute/runtime/NEON/NEFunctions.h"
+#include "support/ToolchainSupport.h"
+
+using namespace arm_compute::utils::cast;
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+namespace
+{
+/** Returns backing tensor of a given tensor
+ *
+ * @param[in] tensor Tensor to extract the backing tensor from
+ *
+ * @return Backing tensor if present else nullptr
+ */
+arm_compute::ITensor *get_backing_tensor(arm_compute::graph2::Tensor *tensor)
+{
+ return ((tensor == nullptr) || (tensor->handle() == nullptr)) ? nullptr : &tensor->handle()->tensor();
+}
+
+/** Create a backend activation layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend activation layer function
+ */
+std::unique_ptr<IFunction> create_activation_layer(ActivationLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON ActivationLayerNode node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const ActivationLayerInfo act_info = node.activation_info();
+
+ // Create function
+ auto func = support::cpp14::make_unique<NEActivationLayer>();
+ func->configure(input, output, act_info);
+
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEActivationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Shape: " << input->info()->tensor_shape()
+ << " Activation function: " << act_info.activation()
+ << " a: " << act_info.a()
+ << " b: " << act_info.b()
+ << " InPlace : " << is_in_place_operation(input, output)
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend batch normalization layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend batch normalization layer function
+ */
+std::unique_ptr<IFunction> create_batch_normalization_layer(BatchNormalizationLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON BatchNormalization node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+
+ // TODO (geopin01) : Var and mean are compulsory, switch function to accept nullptr as beta and/or gamma
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 5);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *mean = get_backing_tensor(node.input(1));
+ ITensor *var = get_backing_tensor(node.input(2));
+ ITensor *beta = get_backing_tensor(node.input(3));
+ ITensor *gamma = get_backing_tensor(node.input(4));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const float epsilon = node.epsilon();
+ const ActivationLayerInfo fused_act = node.fused_activation();
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEBatchNormalizationLayer>();
+ func->configure(input, output, mean, var, beta, gamma, epsilon, fused_act);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEBatchNormalizationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Shape: " << input->info()->tensor_shape()
+ << " Epsilon: " << epsilon << " "
+ << (fused_act.enabled() ? to_string(fused_act.activation()) : "")
+ << " InPlace : " << is_in_place_operation(input, output)
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend convolution layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend convolution layer function
+ */
+std::unique_ptr<IFunction> create_convolution_layer(ConvolutionLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON ConvolutionLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *weights = get_backing_tensor(node.input(1));
+ ITensor *biases = get_backing_tensor(node.input(2));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const PadStrideInfo conv_info = node.convolution_info();
+ const ConvolutionMethod conv_algorithm = node.convolution_method();
+
+ // Create and configure function (we assume that functions have been validated before creation)
+ std::shared_ptr<IMemoryManager> mm = get_memory_manager(ctx, Target::NEON);
+ std::unique_ptr<IFunction> func;
+ std::string func_name;
+ if(conv_algorithm == ConvolutionMethod::DIRECT)
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<NEDirectConvolutionLayer>(std::string("NEDirectConvolutionLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+ else if(conv_algorithm == ConvolutionMethod::GEMM)
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<NEGEMMConvolutionLayer>(std::string("NEGEMMConvolutionLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+ else if(conv_algorithm == ConvolutionMethod::WINOGRAD)
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<NEWinogradLayer>(std::string("NEWinogradLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+ else
+ {
+ std::tie(func, func_name) = create_named_memory_managed_function<NEConvolutionLayer>(std::string("NEConvolutionLayer"), mm,
+ input, weights, biases, output, conv_info);
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+ return func;
+}
+
+/** Create a backend layer depth concatenate function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend depth concatenate layer function
+ */
+std::unique_ptr<arm_compute::IFunction> create_depth_concatenate_layer(DepthConcatenateLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON DepthConcatenate node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Return nullptr if depth concatenate is switched off
+ if(!node.is_enabled())
+ {
+ return nullptr;
+ }
+
+ // Extract IO and info
+ std::vector<arm_compute::ITensor *> inputs;
+ for(unsigned int i = 0; i < node.num_inputs(); ++i)
+ {
+ inputs.push_back(get_backing_tensor(node.input(i)));
+ }
+ ITensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEDepthConcatenateLayer>();
+ func->configure(inputs, output);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEDepthConcatenateLayer"
+ << " Data Type: " << output->info()->data_type()
+ << " Shape: " << output->info()->tensor_shape()
+ << " Num Inputs: " << inputs.size()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend layer depth-wise convolution function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend depth-wise convolution layer function
+ */
+std::unique_ptr<IFunction> create_depthwise_convolution_layer(DepthwiseConvolutionLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON DepthwiseConvolutionLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *weights = get_backing_tensor(node.input(1));
+ ITensor *biases = get_backing_tensor(node.input(2));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const PadStrideInfo conv_info = node.convolution_info();
+ const DepthwiseConvolutionMethod dwc_algorithm = node.depthwise_convolution_method();
+
+ // Create and configure function (we assume that functions have been validated before creation)
+ std::unique_ptr<IFunction> func;
+ std::string func_name;
+ if(dwc_algorithm == DepthwiseConvolutionMethod::OPTIMIZED_3x3)
+ {
+ std::tie(func, func_name) = create_named_function<NEDepthwiseConvolutionLayer3x3>(std::string("NEDepthwiseConvolutionLayer3x3"),
+ input, weights, biases, output, conv_info);
+ }
+ else
+ {
+ std::tie(func, func_name) = create_named_function<NEDepthwiseConvolutionLayer>(std::string("NEDepthwiseConvolutionLayer"),
+ input, weights, biases, output, conv_info);
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+ return func;
+}
+
+/** Create a backend element-wise operation layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend element-wise operation layer function
+ */
+std::unique_ptr<IFunction> create_eltwise_layer(EltwiseLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON EltwiseLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 2);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input1 = get_backing_tensor(node.input(0));
+ ITensor *input2 = get_backing_tensor(node.input(1));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const EltwiseOperation eltwise_op = node.eltwise_operation();
+ ARM_COMPUTE_ERROR_ON(input1 == nullptr);
+ ARM_COMPUTE_ERROR_ON(input2 == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ std::unique_ptr<IFunction> func = nullptr;
+ std::string func_name;
+ if(eltwise_op == EltwiseOperation::ADD)
+ {
+ std::tie(func, func_name) = create_named_function<NEArithmeticAddition>(std::string("NEArithmeticAddition"),
+ input1, input2, output, ConvertPolicy::SATURATE);
+ }
+ else if(eltwise_op == EltwiseOperation::SUB)
+ {
+ std::tie(func, func_name) = create_named_function<NEArithmeticSubtraction>(std::string("NEArithmeticSubtraction"),
+ input1, input2, output, ConvertPolicy::SATURATE);
+ }
+ else if(eltwise_op == EltwiseOperation::MUL)
+ {
+ std::tie(func, func_name) = create_named_function<NEPixelWiseMultiplication>(std::string("NEPixelWiseMultiplication"),
+ input1, input2, output, 1.f,
+ ConvertPolicy::SATURATE, RoundingPolicy::TO_NEAREST_EVEN);
+ }
+ else
+ {
+ ARM_COMPUTE_ERROR("Unsupported element-wise operation!");
+ }
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated " << func_name
+ << " Data Type: " << input1->info()->data_type()
+ << " Shape : " << input1->info()->tensor_shape()
+ << std::endl);
+
+ return func;
+}
+
+/** Create a backend flatten layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend flatten layer function
+ */
+std::unique_ptr<IFunction> create_flatten_layer(FlattenLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON FlattenLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEFlattenLayer>();
+ func->configure(input, output);
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEFlattenLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend fully connected layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend fully connected layer function
+ */
+std::unique_ptr<IFunction> create_fully_connected_layer(FullyConnectedLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON FullyConnectedLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 3);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *weights = get_backing_tensor(node.input(1));
+ ITensor *biases = get_backing_tensor(node.input(2));
+ ITensor *output = get_backing_tensor(node.output(0));
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEFullyConnectedLayer>(get_memory_manager(ctx, Target::NEON));
+ func->configure(input, weights, biases, output);
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(weights == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEFullyConnectedLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Weights shape: " << weights->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend normalization layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend normalization layer function
+ */
+std::unique_ptr<IFunction> create_normalization_layer(NormalizationLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON NormalizationLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const NormalizationLayerInfo norm_info = node.normalization_info();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NENormalizationLayer>(get_memory_manager(ctx, Target::NEON));
+ func->configure(input, output, norm_info);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NENormalizationLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << " Normalization info: " << norm_info.type()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend pooling layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend pooling layer function
+ */
+std::unique_ptr<IFunction> create_pooling_layer(PoolingLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON PoolingLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const PoolingLayerInfo pool_info = node.pooling_info();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEPoolingLayer>();
+ func->configure(input, output, pool_info);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEPoolingLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << " Pooling info: " << pool_info.pool_type()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend reshape layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend reshape layer function
+ */
+std::unique_ptr<IFunction> create_reshape_layer(ReshapeLayerNode &node)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON ReshapeLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NEReshapeLayer>();
+ func->configure(input, output);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NEReshapeLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+
+/** Create a backend softmax layer function
+ *
+ * @param[in] node Node to create the backend function for
+ *
+ * @return Backend softmax layer function
+ */
+std::unique_ptr<IFunction> create_softmax_layer(SoftmaxLayerNode &node, GraphContext &ctx)
+{
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Creating NEON SoftmaxLayer node with ID : " << node.id() << " and Name: " << node.name() << std::endl);
+ ARM_COMPUTE_ERROR_ON(node.num_inputs() != 1);
+ ARM_COMPUTE_ERROR_ON(node.num_outputs() != 1);
+
+ // Extract IO and info
+ ITensor *input = get_backing_tensor(node.input(0));
+ ITensor *output = get_backing_tensor(node.output(0));
+ const float beta = node.beta();
+ ARM_COMPUTE_ERROR_ON(input == nullptr);
+ ARM_COMPUTE_ERROR_ON(output == nullptr);
+
+ // Create and configure function
+ auto func = support::cpp14::make_unique<NESoftmaxLayer>(get_memory_manager(ctx, Target::NEON));
+ func->configure(input, output, beta);
+
+ // Log info
+ ARM_COMPUTE_LOG_GRAPH_INFO("Instantiated NESoftmaxLayer"
+ << " Data Type: " << input->info()->data_type()
+ << " Input shape: " << input->info()->tensor_shape()
+ << " Output shape: " << output->info()->tensor_shape()
+ << std::endl);
+
+ return std::move(func);
+}
+} // namespace
+
+std::unique_ptr<IFunction> NEFunctionFactory::create(INode *node, GraphContext &ctx)
+{
+ if(node == nullptr)
+ {
+ return nullptr;
+ }
+
+ NodeType type = node->type();
+ switch(type)
+ {
+ case NodeType::ActivationLayer:
+ return create_activation_layer(*polymorphic_downcast<ActivationLayerNode *>(node));
+ case NodeType::BatchNormalizationLayer:
+ return create_batch_normalization_layer(*polymorphic_downcast<BatchNormalizationLayerNode *>(node));
+ case NodeType::ConvolutionLayer:
+ return create_convolution_layer(*polymorphic_downcast<ConvolutionLayerNode *>(node), ctx);
+ case NodeType::DepthConcatenateLayer:
+ return create_depth_concatenate_layer(*polymorphic_downcast<DepthConcatenateLayerNode *>(node));
+ case NodeType::DepthwiseConvolutionLayer:
+ return create_depthwise_convolution_layer(*polymorphic_downcast<DepthwiseConvolutionLayerNode *>(node));
+ case NodeType::EltwiseLayer:
+ return create_eltwise_layer(*polymorphic_downcast<EltwiseLayerNode *>(node));
+ case NodeType::FlattenLayer:
+ return create_flatten_layer(*polymorphic_downcast<FlattenLayerNode *>(node));
+ case NodeType::FullyConnectedLayer:
+ return create_fully_connected_layer(*polymorphic_downcast<FullyConnectedLayerNode *>(node), ctx);
+ case NodeType::NormalizationLayer:
+ return create_normalization_layer(*polymorphic_downcast<NormalizationLayerNode *>(node), ctx);
+ case NodeType::PoolingLayer:
+ return create_pooling_layer(*polymorphic_downcast<PoolingLayerNode *>(node));
+ case NodeType::ReshapeLayer:
+ return create_reshape_layer(*polymorphic_downcast<ReshapeLayerNode *>(node));
+ case NodeType::SoftmaxLayer:
+ return create_softmax_layer(*polymorphic_downcast<SoftmaxLayerNode *>(node), ctx);
+ default:
+ return nullptr;
+ }
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/NEON/NESubTensorHandle.cpp b/src/graph2/backends/NEON/NESubTensorHandle.cpp
new file mode 100644
index 0000000000..9b3c9b18d6
--- /dev/null
+++ b/src/graph2/backends/NEON/NESubTensorHandle.cpp
@@ -0,0 +1,70 @@
+/*
+ * 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 "arm_compute/graph2/backends/NEON/NESubTensorHandle.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+NESubTensorHandle::NESubTensorHandle(ITensorHandle *parent_handle, const TensorShape &shape, const Coordinates &coords)
+ : _sub_tensor()
+{
+ ARM_COMPUTE_ERROR_ON(!parent_handle);
+ _sub_tensor = arm_compute::SubTensor(&parent_handle->tensor(), shape, coords);
+}
+
+void NESubTensorHandle::allocate()
+{
+ // noop
+}
+
+const arm_compute::ITensor &NESubTensorHandle::tensor() const
+{
+ return _sub_tensor;
+}
+
+arm_compute::ITensor &NESubTensorHandle::tensor()
+{
+ return _sub_tensor;
+}
+
+void NESubTensorHandle::map(bool blocking)
+{
+ ARM_COMPUTE_UNUSED(blocking);
+}
+
+void NESubTensorHandle::unmap()
+{
+ // noop
+}
+
+bool NESubTensorHandle::is_subtensor() const
+{
+ return true;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/backends/NEON/NETensorHandle.cpp b/src/graph2/backends/NEON/NETensorHandle.cpp
new file mode 100644
index 0000000000..a4af8aaf9b
--- /dev/null
+++ b/src/graph2/backends/NEON/NETensorHandle.cpp
@@ -0,0 +1,68 @@
+/*
+ * 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 "arm_compute/graph2/backends/NEON/NETensorHandle.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace backends
+{
+NETensorHandle::NETensorHandle(const ITensorInfo &info)
+ : _tensor()
+{
+ _tensor.allocator()->init(info);
+}
+
+void NETensorHandle::allocate()
+{
+ _tensor.allocator()->allocate();
+}
+
+const arm_compute::ITensor &NETensorHandle::tensor() const
+{
+ return _tensor;
+}
+
+arm_compute::ITensor &NETensorHandle::tensor()
+{
+ return _tensor;
+}
+
+void NETensorHandle::map(bool blocking)
+{
+ ARM_COMPUTE_UNUSED(blocking);
+}
+
+void NETensorHandle::unmap()
+{
+}
+
+bool NETensorHandle::is_subtensor() const
+{
+ return false;
+}
+} // namespace backends
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/detail/ExecutionHelpers.cpp b/src/graph2/detail/ExecutionHelpers.cpp
new file mode 100644
index 0000000000..a7eba0fec8
--- /dev/null
+++ b/src/graph2/detail/ExecutionHelpers.cpp
@@ -0,0 +1,170 @@
+/*
+ * 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 "arm_compute/graph2/detail/ExecutionHelpers.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/GraphContext.h"
+#include "arm_compute/graph2/GraphManager.h"
+#include "arm_compute/graph2/Tensor.h"
+#include "arm_compute/graph2/backends/BackendRegistry.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace detail
+{
+void default_initialize_backends()
+{
+ for(const auto &backend : backends::BackendRegistry::get().backends())
+ {
+ backend.second->initialize_backend();
+ }
+}
+
+void configure_all_tensors(Graph &g)
+{
+ auto &tensors = g.tensors();
+
+ for(auto &tensor : tensors)
+ {
+ if(tensor)
+ {
+ Target target = tensor->desc().target;
+ auto backend = backends::BackendRegistry::get().find_backend(target);
+ ARM_COMPUTE_ERROR_ON_MSG(!backend, "Requested backend doesn't exist!");
+ auto handle = backend->create_tensor(*tensor);
+ ARM_COMPUTE_ERROR_ON_MSG(!backend, "Couldn't create backend handle!");
+ tensor->set_handle(std::move(handle));
+ }
+ }
+}
+
+void allocate_all_tensors(Graph &g)
+{
+ auto &tensors = g.tensors();
+
+ for(auto &tensor : tensors)
+ {
+ if(tensor && !tensor->bound_edges().empty())
+ {
+ ARM_COMPUTE_ERROR_ON_MSG(!tensor->handle(), "Tensor handle is not configured!");
+ tensor->handle()->allocate();
+ }
+ }
+}
+
+ExecutionWorkload configure_all_nodes(Graph &g, GraphContext &ctx)
+{
+ ExecutionWorkload workload;
+ auto &nodes = g.nodes();
+
+ // Create tasks
+ for(auto &node : nodes)
+ {
+ if(node != nullptr)
+ {
+ Target assigned_target = node->assigned_target();
+ auto backend = backends::BackendRegistry::get().find_backend(assigned_target);
+ ARM_COMPUTE_ERROR_ON_MSG(!backend, "Requested backend doesn't exist!");
+ auto func = backend->configure_node(*node, ctx);
+ if(func != nullptr)
+ {
+ ExecutionTask task;
+ task.task = std::move(func);
+ task.node = node.get();
+ workload.tasks.push_back(std::move(task));
+ }
+ }
+ }
+
+ // Add inputs and outputs
+ for(auto &node : nodes)
+ {
+ if(node != nullptr && node->type() == NodeType::Input)
+ {
+ workload.inputs.push_back(node->output(0));
+ }
+
+ if(node != nullptr && node->type() == NodeType::Output)
+ {
+ workload.outputs.push_back(node->input(0));
+ continue;
+ }
+ }
+
+ return workload;
+}
+
+void call_tensor_accessor(Tensor *tensor)
+{
+ ARM_COMPUTE_ERROR_ON(!tensor);
+ tensor->call_accessor();
+}
+
+void call_all_const_node_accessors(Graph &g)
+{
+ auto &nodes = g.nodes();
+
+ for(auto &node : nodes)
+ {
+ if(node != nullptr && node->type() == NodeType::Const)
+ {
+ call_tensor_accessor(node->output(0));
+ }
+ }
+}
+
+void call_all_input_node_accessors(ExecutionWorkload &workload)
+{
+ for(auto &input : workload.inputs)
+ {
+ if(input != nullptr)
+ {
+ input->call_accessor();
+ }
+ }
+}
+
+void call_all_tasks(ExecutionWorkload &workload)
+{
+ for(auto &task : workload.tasks)
+ {
+ task();
+ }
+}
+
+void call_all_output_node_accessors(ExecutionWorkload &workload)
+{
+ for(auto &output : workload.outputs)
+ {
+ if(output != nullptr)
+ {
+ output->call_accessor();
+ }
+ }
+}
+} // namespace detail
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/frontend/Stream.cpp b/src/graph2/frontend/Stream.cpp
new file mode 100644
index 0000000000..076b9ac11f
--- /dev/null
+++ b/src/graph2/frontend/Stream.cpp
@@ -0,0 +1,70 @@
+/*
+ * 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 "arm_compute/graph2/frontend/Stream.h"
+
+#include "arm_compute/graph2/Utils.h"
+#include "arm_compute/graph2/frontend/ILayer.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace frontend
+{
+Stream::Stream(size_t id, std::string name)
+ : _manager(), _ctx(), _g(id, std::move(name))
+{
+}
+
+void Stream::finalize(Target target, bool enable_tuning, bool enable_memory_management)
+{
+ PassManager pm = create_default_pass_manager();
+ _ctx.enable_tuning(enable_tuning);
+ _ctx.enable_memory_managenent(enable_memory_management);
+ _manager.finalize_graph(_g, _ctx, pm, target);
+}
+
+void Stream::run()
+{
+ _manager.execute_graph(_g);
+}
+
+void Stream::add_layer(ILayer &layer)
+{
+ auto nid = layer.create_layer(*this);
+ _tail_node = nid;
+}
+
+const Graph &Stream::graph() const
+{
+ return _g;
+}
+
+Graph &Stream::graph()
+{
+ return _g;
+}
+} // namespace frontend
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/frontend/SubStream.cpp b/src/graph2/frontend/SubStream.cpp
new file mode 100644
index 0000000000..e6fa605ad1
--- /dev/null
+++ b/src/graph2/frontend/SubStream.cpp
@@ -0,0 +1,59 @@
+/*
+ * 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 "arm_compute/graph2/frontend/SubStream.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/frontend/ILayer.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace frontend
+{
+SubStream::SubStream(IStream &s)
+ : _s(s)
+{
+ _hints = s.hints();
+ _tail_node = s.tail_node();
+}
+
+void SubStream::add_layer(ILayer &layer)
+{
+ auto nid = layer.create_layer(*this);
+ _tail_node = nid;
+}
+
+const Graph &SubStream::graph() const
+{
+ return _s.graph();
+}
+
+Graph &SubStream::graph()
+{
+ return _s.graph();
+}
+} // namespace frontend
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/mutators/DepthConcatSubTensorMutator.cpp b/src/graph2/mutators/DepthConcatSubTensorMutator.cpp
new file mode 100644
index 0000000000..cc8de6bb1b
--- /dev/null
+++ b/src/graph2/mutators/DepthConcatSubTensorMutator.cpp
@@ -0,0 +1,86 @@
+/*
+ * 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 "arm_compute/graph2/mutators/DepthConcatSubTensorMutator.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/backends/BackendRegistry.h"
+#include "arm_compute/graph2/nodes/DepthConcatenateLayerNode.h"
+
+#include "arm_compute/core/utils/misc/Cast.h"
+#include "arm_compute/core/utils/misc/Iterable.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+const char *DepthConcatSubTensorMutator::name()
+{
+ return "DepthConcatSubTensorMutator";
+}
+
+void DepthConcatSubTensorMutator::mutate(Graph &g)
+{
+ // Should be in reverse order of execution
+ for(auto &node : arm_compute::utils::iterable::reverse_iterate(g.nodes()))
+ {
+ if(node && node->type() == NodeType::DepthConcatenateLayer && node->output(0) != nullptr)
+ {
+ // Get output tensor
+ auto output_tensor = node->output(0);
+
+ // Check that all tensor have the same target and valid inputs
+ bool is_valid = std::all_of(node->input_edges().cbegin(), node->input_edges().cend(),
+ [&](const EdgeID & eid)
+ {
+ return (g.edge(eid) != nullptr) && (g.edge(eid)->tensor() != nullptr) && (g.edge(eid)->tensor()->desc().target == output_tensor->desc().target);
+ });
+
+ // Create subtensors
+ if(is_valid && backends::BackendRegistry::get().find_backend(output_tensor->desc().target) != nullptr)
+ {
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Using sub-tensors for the node with ID : "
+ << node->id() << " and name : " << node->name() << std::endl);
+ // Create sub-tensor handles
+ unsigned depth = 0;
+ for(unsigned int i = 0; i < node->input_edges().size(); ++i)
+ {
+ auto input_tensor = node->input(i);
+ const auto input_shape = input_tensor->desc().shape;
+
+ auto backend = backends::BackendRegistry::get().find_backend(input_tensor->desc().target);
+ auto handle = backend->create_subtensor(output_tensor->handle(), input_shape, Coordinates(0, 0, depth));
+ input_tensor->set_handle(std::move(handle));
+
+ depth += input_shape.z();
+ }
+
+ auto *dc_node = arm_compute::utils::cast::polymorphic_downcast<DepthConcatenateLayerNode *>(node.get());
+ dc_node->set_enabled(false);
+ }
+ }
+ }
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/mutators/InPlaceOperationMutator.cpp b/src/graph2/mutators/InPlaceOperationMutator.cpp
new file mode 100644
index 0000000000..bb13e98999
--- /dev/null
+++ b/src/graph2/mutators/InPlaceOperationMutator.cpp
@@ -0,0 +1,63 @@
+/*
+ * 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 "arm_compute/graph2/mutators/InPlaceOperationMutator.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Logger.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+const char *InPlaceOperationMutator::name()
+{
+ return "InPlaceOperationMutator";
+}
+
+void InPlaceOperationMutator::mutate(Graph &g)
+{
+ std::set<NodeType> in_place_nodes = { NodeType::BatchNormalizationLayer, NodeType::ActivationLayer };
+
+ // Not interested in the order of nodes
+ for(auto &node : g.nodes())
+ {
+ if(node && in_place_nodes.find(node->type()) != std::end(in_place_nodes))
+ {
+ // Get input edge
+ Edge *input_edge = node->input_edge(0);
+
+ // Check if parent has a single output if yes then force in place calculation else not
+ if((input_edge != nullptr) && (input_edge->producer() != nullptr) && (input_edge->producer()->output_edges().size() == 1))
+ {
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Switching to in-place computation for the node with ID : "
+ << node->id() << " and name : " << node->name() << std::endl);
+ // Update output
+ auto tensor = input_edge->tensor();
+ node->set_output_tensor(tensor->id(), 0);
+ }
+ }
+ }
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/mutators/NodeFusionMutator.cpp b/src/graph2/mutators/NodeFusionMutator.cpp
new file mode 100644
index 0000000000..d0ab3e7e6b
--- /dev/null
+++ b/src/graph2/mutators/NodeFusionMutator.cpp
@@ -0,0 +1,96 @@
+/*
+ * 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 "arm_compute/graph2/mutators/NodeFusionMutator.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Logger.h"
+#include "arm_compute/graph2/nodes/Nodes.h"
+
+#include "arm_compute/core/utils/misc/Cast.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+namespace detail
+{
+void fuse_batch_norm_with_activation(Graph &g)
+{
+ // Not interested in the order of nodes
+ for(auto &node : g.nodes())
+ {
+ // Check if the node is batch norm and not a branching node
+ if(node && node->type() == NodeType::BatchNormalizationLayer && node->output_edges().size() == 1)
+ {
+ auto output_edge_id = *node->output_edges().begin();
+ auto output_edge = g.edge(output_edge_id);
+ // Check if following node is an activation layer node
+ if((output_edge != nullptr) && (output_edge->consumer() != nullptr) && (output_edge->consumer()->type() == NodeType::ActivationLayer))
+ {
+ ARM_COMPUTE_LOG_GRAPH_VERBOSE("Fusing Batch Normalization node with ID : " << output_edge->producer_id()
+ << " with Activation Layer node with ID : " << output_edge->consumer_id() << std::endl);
+
+ auto *bn_node = arm_compute::utils::cast::polymorphic_downcast<BatchNormalizationLayerNode *>(output_edge->producer());
+ auto *act_node = arm_compute::utils::cast::polymorphic_downcast<ActivationLayerNode *>(output_edge->consumer());
+
+ // Get driving nodes of activation node
+ std::vector<NodeIdxPair> act_driving_nodes;
+ for(auto &act_output_edge_id : act_node->output_edges())
+ {
+ auto act_output_edge = g.edge(act_output_edge_id);
+ if(act_output_edge != nullptr)
+ {
+ ARM_COMPUTE_ERROR_ON(act_output_edge->consumer() == nullptr);
+ act_driving_nodes.push_back({ act_output_edge->consumer_id(), act_output_edge->consumer_idx() });
+ }
+ }
+
+ // Set activation info to batch normalization
+ bn_node->set_fused_activation(act_node->activation_info());
+
+ // Remove activation node
+ g.remove_node(act_node->id());
+
+ // Update batch normalization node outputs
+ for(auto &driving_node : act_driving_nodes)
+ {
+ g.add_connection(bn_node->id(), 0, driving_node.node_id, driving_node.index);
+ }
+ }
+ }
+ }
+}
+} // namespace detail
+
+const char *NodeFusionMutator::name()
+{
+ return "NodeFusionMutator";
+}
+
+void NodeFusionMutator::mutate(Graph &g)
+{
+ detail::fuse_batch_norm_with_activation(g);
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/nodes/ActivationLayerNode.cpp b/src/graph2/nodes/ActivationLayerNode.cpp
new file mode 100644
index 0000000000..c7c36e9bbd
--- /dev/null
+++ b/src/graph2/nodes/ActivationLayerNode.cpp
@@ -0,0 +1,83 @@
+/*
+ * 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 "arm_compute/graph2/nodes/ActivationLayerNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+ActivationLayerNode::ActivationLayerNode(ActivationLayerInfo info)
+ : _info(info)
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+ActivationLayerInfo ActivationLayerNode::activation_info() const
+{
+ return _info;
+}
+
+bool ActivationLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor ActivationLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ return src->desc();
+}
+
+Status ActivationLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType ActivationLayerNode::type() const
+{
+ return NodeType::ActivationLayer;
+}
+
+void ActivationLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/BatchNormalizationLayerNode.cpp b/src/graph2/nodes/BatchNormalizationLayerNode.cpp
new file mode 100644
index 0000000000..b9f634210c
--- /dev/null
+++ b/src/graph2/nodes/BatchNormalizationLayerNode.cpp
@@ -0,0 +1,94 @@
+/*
+ * 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 "arm_compute/graph2/nodes/BatchNormalizationLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+BatchNormalizationLayerNode::BatchNormalizationLayerNode(float epsilon, ActivationLayerInfo fused_activation)
+ : _epsilon(epsilon), _fused_activation(fused_activation)
+{
+ _input_edges.resize(5, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+float BatchNormalizationLayerNode::epsilon() const
+{
+ return _epsilon;
+}
+
+ActivationLayerInfo BatchNormalizationLayerNode::fused_activation() const
+{
+ return _fused_activation;
+}
+
+void BatchNormalizationLayerNode::set_fused_activation(ActivationLayerInfo fused_activation)
+{
+ _fused_activation = fused_activation;
+}
+
+bool BatchNormalizationLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor BatchNormalizationLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ return src->desc();
+}
+
+Status BatchNormalizationLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType BatchNormalizationLayerNode::type() const
+{
+ return NodeType::BatchNormalizationLayer;
+}
+
+void BatchNormalizationLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/ConstNode.cpp b/src/graph2/nodes/ConstNode.cpp
new file mode 100644
index 0000000000..5bd6a8180c
--- /dev/null
+++ b/src/graph2/nodes/ConstNode.cpp
@@ -0,0 +1,72 @@
+/*
+ * 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 "arm_compute/graph2/nodes/ConstNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+ConstNode::ConstNode(TensorDescriptor desc)
+ : _desc(desc)
+{
+ _outputs.resize(1, NullTensorID);
+}
+
+bool ConstNode::forward_descriptors()
+{
+ if(output_id(0) != NullTensorID)
+ {
+ Tensor *t = output(0);
+ ARM_COMPUTE_ERROR_ON(t == nullptr);
+ t->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor ConstNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ return _desc;
+}
+
+Status ConstNode::validate()
+{
+ return Status{};
+}
+
+NodeType ConstNode::type() const
+{
+ return NodeType::Const;
+}
+
+void ConstNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/nodes/ConvolutionLayerNode.cpp b/src/graph2/nodes/ConvolutionLayerNode.cpp
new file mode 100644
index 0000000000..499b3520b2
--- /dev/null
+++ b/src/graph2/nodes/ConvolutionLayerNode.cpp
@@ -0,0 +1,111 @@
+/*
+ * 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 "arm_compute/graph2/nodes/ConvolutionLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+ConvolutionLayerNode::ConvolutionLayerNode(PadStrideInfo info, ConvolutionMethod method)
+ : _info(std::move(info)), _method(method)
+{
+ _input_edges.resize(3, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+void ConvolutionLayerNode::set_convolution_method(ConvolutionMethod method)
+{
+ _method = method;
+}
+
+ConvolutionMethod ConvolutionLayerNode::convolution_method() const
+{
+ return _method;
+}
+
+PadStrideInfo ConvolutionLayerNode::convolution_info() const
+{
+ return _info;
+}
+
+TensorShape ConvolutionLayerNode::compute_output_shape(TensorShape input_shape, TensorShape weights_shape, PadStrideInfo info)
+{
+ unsigned int output_width = 0;
+ unsigned int output_height = 0;
+ std::tie(output_width, output_height) = scaled_dimensions(input_shape.x(), input_shape.y(), weights_shape.x(), weights_shape.y(), info);
+
+ TensorShape output_shape{ input_shape };
+ output_shape.set(0, output_width);
+ output_shape.set(1, output_height);
+ output_shape.set(2, weights_shape[3]);
+
+ return output_shape;
+}
+
+bool ConvolutionLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (input_id(1) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor ConvolutionLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ const Tensor *src = input(0);
+ const Tensor *weights = input(1);
+
+ ARM_COMPUTE_ERROR_ON(src == nullptr || weights == nullptr);
+
+ TensorDescriptor output_info = src->desc();
+ TensorShape output_shape = compute_output_shape(src->desc().shape, weights->desc().shape, _info);
+ output_info.shape = output_shape;
+ return output_info;
+}
+
+Status ConvolutionLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType ConvolutionLayerNode::type() const
+{
+ return NodeType::ConvolutionLayer;
+}
+
+void ConvolutionLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/DepthConcatenateLayerNode.cpp b/src/graph2/nodes/DepthConcatenateLayerNode.cpp
new file mode 100644
index 0000000000..dcd66517b6
--- /dev/null
+++ b/src/graph2/nodes/DepthConcatenateLayerNode.cpp
@@ -0,0 +1,133 @@
+/*
+ * 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 "arm_compute/graph2/nodes/DepthConcatenateLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+DepthConcatenateLayerNode::DepthConcatenateLayerNode(unsigned int total_nodes)
+ : _total_nodes(total_nodes), _is_enabled(true)
+{
+ _input_edges.resize(total_nodes, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+void DepthConcatenateLayerNode::set_enabled(bool is_enabled)
+{
+ _is_enabled = is_enabled;
+}
+
+bool DepthConcatenateLayerNode::is_enabled() const
+{
+ return _is_enabled;
+}
+
+TensorShape DepthConcatenateLayerNode::compute_output_shape(const std::vector<TensorShape> &input_shapes)
+{
+ ARM_COMPUTE_ERROR_ON(input_shapes.size() == 0);
+
+ TensorShape output_shape = input_shapes[0];
+
+ size_t max_x = 0;
+ size_t max_y = 0;
+ size_t depth = 0;
+
+ for(const auto &shape : input_shapes)
+ {
+ max_x = std::max(shape.x(), max_x);
+ max_y = std::max(shape.y(), max_y);
+ depth += shape.z();
+ }
+
+ output_shape.set(0, max_x);
+ output_shape.set(1, max_y);
+ output_shape.set(2, depth);
+
+ return output_shape;
+}
+
+bool DepthConcatenateLayerNode::forward_descriptors()
+{
+ if(_outputs[0] != NullTensorID)
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor DepthConcatenateLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ // Check if all input tensors are set
+ bool are_all_inputs_set = std::all_of(std::begin(_input_edges), std::end(_input_edges), [](const EdgeID & eid)
+ {
+ return eid != EmptyEdgeID;
+ });
+
+ TensorDescriptor output_info = {};
+
+ if(are_all_inputs_set)
+ {
+ std::vector<TensorShape> inputs_shapes;
+ for(unsigned int i = 0; i < _input_edges.size(); ++i)
+ {
+ const Tensor *t = _graph->tensor(input_id(i));
+ ARM_COMPUTE_ERROR_ON(t == nullptr);
+ inputs_shapes.push_back(t->desc().shape);
+ }
+ output_info = input(0)->desc();
+ TensorShape output_shape = compute_output_shape(inputs_shapes);
+ output_info.shape = output_shape;
+ }
+
+ return output_info;
+}
+
+Status DepthConcatenateLayerNode::validate()
+{
+ ARM_COMPUTE_UNUSED(_total_nodes);
+ return Status{};
+}
+
+NodeType DepthConcatenateLayerNode::type() const
+{
+ return NodeType::DepthConcatenateLayer;
+}
+
+void DepthConcatenateLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/DepthwiseConvolutionLayerNode.cpp b/src/graph2/nodes/DepthwiseConvolutionLayerNode.cpp
new file mode 100644
index 0000000000..b030e8b7ca
--- /dev/null
+++ b/src/graph2/nodes/DepthwiseConvolutionLayerNode.cpp
@@ -0,0 +1,110 @@
+/*
+ * 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 "arm_compute/graph2/nodes/DepthwiseConvolutionLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+DepthwiseConvolutionLayerNode::DepthwiseConvolutionLayerNode(PadStrideInfo info, DepthwiseConvolutionMethod method)
+ : _info(std::move(info)), _method(method)
+{
+ _input_edges.resize(3, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+void DepthwiseConvolutionLayerNode::set_depthwise_convolution_method(DepthwiseConvolutionMethod method)
+{
+ _method = method;
+}
+
+DepthwiseConvolutionMethod DepthwiseConvolutionLayerNode::depthwise_convolution_method() const
+{
+ return _method;
+}
+
+PadStrideInfo DepthwiseConvolutionLayerNode::convolution_info() const
+{
+ return _info;
+}
+
+TensorShape DepthwiseConvolutionLayerNode::compute_output_shape(TensorShape input_shape, TensorShape weights_shape, PadStrideInfo info)
+{
+ unsigned int output_width = 0;
+ unsigned int output_height = 0;
+ std::tie(output_width, output_height) = scaled_dimensions(input_shape.x(), input_shape.y(), weights_shape.x(), weights_shape.y(), info);
+
+ TensorShape output_shape{ input_shape };
+ output_shape.set(0, output_width);
+ output_shape.set(1, output_height);
+
+ return output_shape;
+}
+
+bool DepthwiseConvolutionLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (input_id(1) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor DepthwiseConvolutionLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ const Tensor *src = input(0);
+ const Tensor *weights = input(1);
+
+ ARM_COMPUTE_ERROR_ON(src == nullptr || weights == nullptr);
+
+ TensorDescriptor output_info = src->desc();
+ TensorShape output_shape = compute_output_shape(src->desc().shape, weights->desc().shape, _info);
+ output_info.shape = output_shape;
+ return output_info;
+}
+
+Status DepthwiseConvolutionLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType DepthwiseConvolutionLayerNode::type() const
+{
+ return NodeType::DepthwiseConvolutionLayer;
+}
+
+void DepthwiseConvolutionLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/EltwiseLayerNode.cpp b/src/graph2/nodes/EltwiseLayerNode.cpp
new file mode 100644
index 0000000000..149d926d29
--- /dev/null
+++ b/src/graph2/nodes/EltwiseLayerNode.cpp
@@ -0,0 +1,83 @@
+/*
+ * 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 "arm_compute/graph2/nodes/EltwiseLayerNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+EltwiseLayerNode::EltwiseLayerNode(EltwiseOperation op)
+ : _op(op)
+{
+ _input_edges.resize(2, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+EltwiseOperation EltwiseLayerNode::eltwise_operation() const
+{
+ return _op;
+}
+
+bool EltwiseLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor EltwiseLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_UNUSED(_op);
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ return src->desc();
+}
+
+Status EltwiseLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType EltwiseLayerNode::type() const
+{
+ return NodeType::EltwiseLayer;
+}
+
+void EltwiseLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/nodes/FlattenLayerNode.cpp b/src/graph2/nodes/FlattenLayerNode.cpp
new file mode 100644
index 0000000000..7c4059f3ed
--- /dev/null
+++ b/src/graph2/nodes/FlattenLayerNode.cpp
@@ -0,0 +1,80 @@
+/*
+ * 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 "arm_compute/graph2/nodes/FlattenLayerNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+FlattenLayerNode::FlattenLayerNode()
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+bool FlattenLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor FlattenLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ TensorDescriptor output_desc = src->desc();
+ output_desc.shape.collapse(src->desc().shape.num_dimensions());
+
+ return output_desc;
+}
+
+Status FlattenLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType FlattenLayerNode::type() const
+{
+ return NodeType::FlattenLayer;
+}
+
+void FlattenLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/FullyConnectedLayer.cpp b/src/graph2/nodes/FullyConnectedLayer.cpp
new file mode 100644
index 0000000000..195adc40fe
--- /dev/null
+++ b/src/graph2/nodes/FullyConnectedLayer.cpp
@@ -0,0 +1,107 @@
+/*
+ * 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 "arm_compute/graph2/nodes/FullyConnectedLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+FullyConnectedLayerNode::FullyConnectedLayerNode(unsigned int num_outputs)
+ : _num_outputs(num_outputs)
+{
+ _input_edges.resize(3, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+TensorShape FullyConnectedLayerNode::compute_weights_shape(TensorShape input_shape, unsigned int num_outputs)
+{
+ unsigned int num_weights = 1;
+ unsigned int num_dimensions = input_shape.num_dimensions();
+ // Ignore the batch dimension if there is one:
+ if(num_dimensions == 2 || num_dimensions == 4)
+ {
+ num_dimensions--;
+ }
+ for(unsigned int i = 0; i < num_dimensions; i++)
+ {
+ num_weights *= input_shape[i];
+ }
+ return TensorShape(num_weights, num_outputs);
+}
+
+TensorShape FullyConnectedLayerNode::compute_output_shape(TensorShape input_shape, unsigned int num_outputs)
+{
+ // Note: Only 1D batch space is supported at the moment
+ unsigned int batches = input_shape[1];
+ if(input_shape.num_dimensions() > 2)
+ {
+ batches = input_shape[3];
+ }
+ return TensorShape(num_outputs, batches);
+}
+
+bool FullyConnectedLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor FullyConnectedLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ TensorDescriptor output_info = src->desc();
+ TensorShape output_shape = compute_output_shape(src->desc().shape, _num_outputs);
+ output_info.shape = output_shape;
+ return output_info;
+}
+
+Status FullyConnectedLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType FullyConnectedLayerNode::type() const
+{
+ return NodeType::FullyConnectedLayer;
+}
+
+void FullyConnectedLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/InputNode.cpp b/src/graph2/nodes/InputNode.cpp
new file mode 100644
index 0000000000..84cce2acdb
--- /dev/null
+++ b/src/graph2/nodes/InputNode.cpp
@@ -0,0 +1,72 @@
+/*
+ * 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 "arm_compute/graph2/nodes/InputNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+InputNode::InputNode(TensorDescriptor desc)
+ : _desc(desc)
+{
+ _outputs.resize(1, NullTensorID);
+}
+
+bool InputNode::forward_descriptors()
+{
+ if(output_id(0) != NullTensorID)
+ {
+ Tensor *t = output(0);
+ ARM_COMPUTE_ERROR_ON(t == nullptr);
+ t->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor InputNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ return _desc;
+}
+
+Status InputNode::validate()
+{
+ return Status{};
+}
+
+NodeType InputNode::type() const
+{
+ return NodeType::Input;
+}
+
+void InputNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/nodes/NormalizationLayerNode.cpp b/src/graph2/nodes/NormalizationLayerNode.cpp
new file mode 100644
index 0000000000..a394879a3e
--- /dev/null
+++ b/src/graph2/nodes/NormalizationLayerNode.cpp
@@ -0,0 +1,84 @@
+/*
+ * 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 "arm_compute/graph2/nodes/NormalizationLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+NormalizationLayerNode::NormalizationLayerNode(NormalizationLayerInfo norm_info)
+ : _info(norm_info)
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+NormalizationLayerInfo NormalizationLayerNode::normalization_info() const
+{
+ return _info;
+}
+
+bool NormalizationLayerNode::forward_descriptors()
+{
+ if(input_id(0) != NullTensorID && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor NormalizationLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ return src->desc();
+}
+
+Status NormalizationLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType NormalizationLayerNode::type() const
+{
+ return NodeType::NormalizationLayer;
+}
+
+void NormalizationLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/OutputNode.cpp b/src/graph2/nodes/OutputNode.cpp
new file mode 100644
index 0000000000..1daebb1cc8
--- /dev/null
+++ b/src/graph2/nodes/OutputNode.cpp
@@ -0,0 +1,66 @@
+/*
+ * 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 "arm_compute/graph2/nodes/OutputNode.h"
+
+#include "arm_compute/core/Error.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+#include "arm_compute/graph2/Tensor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+OutputNode::OutputNode()
+{
+ _input_edges.resize(1, EmptyEdgeID);
+}
+
+bool OutputNode::forward_descriptors()
+{
+ return true;
+}
+
+TensorDescriptor OutputNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ return TensorDescriptor();
+}
+
+Status OutputNode::validate()
+{
+ return Status{};
+}
+
+NodeType OutputNode::type() const
+{
+ return NodeType::Output;
+}
+
+void OutputNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/graph2/nodes/PoolingLayerNode.cpp b/src/graph2/nodes/PoolingLayerNode.cpp
new file mode 100644
index 0000000000..2c2cf5387a
--- /dev/null
+++ b/src/graph2/nodes/PoolingLayerNode.cpp
@@ -0,0 +1,103 @@
+/*
+ * 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 "arm_compute/graph2/nodes/PoolingLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+PoolingLayerNode::PoolingLayerNode(PoolingLayerInfo pool_info)
+ : _info(std::move(pool_info))
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+PoolingLayerInfo PoolingLayerNode::pooling_info() const
+{
+ return _info;
+}
+
+TensorShape PoolingLayerNode::compute_output_shape(TensorShape input_shape, PoolingLayerInfo info)
+{
+ const int pool_size_x = info.is_global_pooling() ? input_shape.x() : info.pool_size().width;
+ const int pool_size_y = info.is_global_pooling() ? input_shape.y() : info.pool_size().height;
+
+ unsigned int pooled_width = 0;
+ unsigned int pooled_height = 0;
+ std::tie(pooled_width, pooled_height) = scaled_dimensions(input_shape.x(), input_shape.y(), pool_size_x, pool_size_y, info.pad_stride_info());
+
+ TensorShape output_shape{ input_shape };
+ output_shape.set(0, pooled_width);
+ output_shape.set(1, pooled_height);
+
+ return output_shape;
+}
+
+bool PoolingLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor PoolingLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ TensorDescriptor output_info = src->desc();
+ TensorShape output_shape = compute_output_shape(src->desc().shape, _info);
+ output_info.shape = output_shape;
+ return output_info;
+}
+
+Status PoolingLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType PoolingLayerNode::type() const
+{
+ return NodeType::PoolingLayer;
+}
+
+void PoolingLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/ReshapeLayer.cpp b/src/graph2/nodes/ReshapeLayer.cpp
new file mode 100644
index 0000000000..6280eea75c
--- /dev/null
+++ b/src/graph2/nodes/ReshapeLayer.cpp
@@ -0,0 +1,81 @@
+/*
+ * 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 "arm_compute/graph2/nodes/ReshapeLayerNode.h"
+
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+ReshapeLayerNode::ReshapeLayerNode(TensorShape shape)
+ : _shape(shape)
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+bool ReshapeLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor ReshapeLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ TensorDescriptor output_desc = src->desc();
+ output_desc.shape = _shape;
+
+ return output_desc;
+}
+
+Status ReshapeLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType ReshapeLayerNode::type() const
+{
+ return NodeType::ReshapeLayer;
+}
+
+void ReshapeLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/nodes/SoftmaxLayerNode.cpp b/src/graph2/nodes/SoftmaxLayerNode.cpp
new file mode 100644
index 0000000000..83bc978981
--- /dev/null
+++ b/src/graph2/nodes/SoftmaxLayerNode.cpp
@@ -0,0 +1,84 @@
+/*
+ * 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 "arm_compute/graph2/nodes/SoftmaxLayerNode.h"
+
+#include "arm_compute/core/Utils.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/INodeVisitor.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+SoftmaxLayerNode::SoftmaxLayerNode(float beta)
+ : _beta(beta)
+{
+ _input_edges.resize(1, EmptyEdgeID);
+ _outputs.resize(1, NullTensorID);
+}
+
+float SoftmaxLayerNode::beta() const
+{
+ return _beta;
+}
+
+bool SoftmaxLayerNode::forward_descriptors()
+{
+ if((input_id(0) != NullTensorID) && (output_id(0) != NullTensorID))
+ {
+ Tensor *dst = output(0);
+ ARM_COMPUTE_ERROR_ON(dst == nullptr);
+ dst->desc() = configure_output(0);
+ return true;
+ }
+ return false;
+}
+
+TensorDescriptor SoftmaxLayerNode::configure_output(size_t idx) const
+{
+ ARM_COMPUTE_UNUSED(idx);
+ ARM_COMPUTE_ERROR_ON(idx >= _outputs.size());
+
+ const Tensor *src = input(0);
+ ARM_COMPUTE_ERROR_ON(src == nullptr);
+
+ return src->desc();
+}
+
+Status SoftmaxLayerNode::validate()
+{
+ return Status{};
+}
+
+NodeType SoftmaxLayerNode::type() const
+{
+ return NodeType::SoftmaxLayer;
+}
+
+void SoftmaxLayerNode::accept(INodeVisitor &v)
+{
+ v.visit(*this);
+}
+} // namespace graph2
+} // namespace arm_compute \ No newline at end of file
diff --git a/src/graph2/printers/DotGraphPrinter.cpp b/src/graph2/printers/DotGraphPrinter.cpp
new file mode 100644
index 0000000000..04987eebe0
--- /dev/null
+++ b/src/graph2/printers/DotGraphPrinter.cpp
@@ -0,0 +1,173 @@
+/*
+ * 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 "arm_compute/graph2/printers/DotGraphPrinter.h"
+
+#include "arm_compute/core/Error.h"
+#include "arm_compute/graph2/Graph.h"
+#include "arm_compute/graph2/Tensor.h"
+#include "arm_compute/graph2/TypePrinter.h"
+#include "arm_compute/graph2/nodes/Nodes.h"
+
+namespace arm_compute
+{
+namespace graph2
+{
+void DotGraphVisitor::visit(ActivationLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.activation_info().activation();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(BatchNormalizationLayerNode &n)
+{
+ std::stringstream ss;
+ ss << (n.fused_activation().enabled() ? to_string(n.fused_activation().activation()) : "");
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(ConvolutionLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.convolution_method();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(DepthConcatenateLayerNode &n)
+{
+ std::stringstream ss;
+ ss << "Enabled: " << n.is_enabled();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(DepthwiseConvolutionLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.depthwise_convolution_method();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(EltwiseLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.eltwise_operation();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(NormalizationLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.normalization_info().type();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::visit(PoolingLayerNode &n)
+{
+ std::stringstream ss;
+ ss << n.pooling_info().pool_type();
+ ss << R"( \n )";
+ ss << n.pooling_info().pool_size();
+ ss << R"( \n )";
+ ss << n.pooling_info().pad_stride_info();
+ _info = ss.str();
+}
+
+void DotGraphVisitor::default_visit()
+{
+ _info.clear();
+}
+
+const std::string &DotGraphVisitor::info() const
+{
+ return _info;
+}
+
+void DotGraphPrinter::print(const Graph &g, std::ostream &os)
+{
+ // Print header
+ print_header(g, os);
+
+ // Print nodes
+ print_nodes(g, os);
+
+ // Print edges
+ print_edges(g, os);
+
+ // Print footer
+ print_footer(g, os);
+}
+
+void DotGraphPrinter::print_header(const Graph &g, std::ostream &os)
+{
+ // Print graph name
+ std::string graph_name = (g.name().empty()) ? "Graph" : g.name();
+ os << "digraph " << graph_name << "{\n";
+}
+
+void DotGraphPrinter::print_footer(const Graph &g, std::ostream &os)
+{
+ ARM_COMPUTE_UNUSED(g);
+ os << "}\n";
+}
+
+void DotGraphPrinter::print_nodes(const Graph &g, std::ostream &os)
+{
+ for(const auto &n : g.nodes())
+ {
+ if(n)
+ {
+ // Output node id
+ std::string node_id = std::string("n") + support::cpp11::to_string(n->id());
+ os << node_id << " ";
+
+ // Output label
+ n->accept(_dot_node_visitor);
+
+ std::string name = n->name().empty() ? node_id : n->name();
+ auto node_description = _dot_node_visitor.info();
+
+ os << R"([label = ")" << name << R"( \n )" << n->assigned_target() << R"( \n )" << node_description << R"("])";
+ os << ";\n";
+ }
+ }
+}
+
+void DotGraphPrinter::print_edges(const Graph &g, std::ostream &os)
+{
+ for(const auto &e : g.edges())
+ {
+ if(e)
+ {
+ std::string source_node_id = std::string("n") + support::cpp11::to_string(e->producer_id());
+ std::string sink_node_id = std::string("n") + support::cpp11::to_string(e->consumer_id());
+ os << source_node_id << " -> " << sink_node_id << " ";
+ const Tensor *t = e->tensor();
+ ARM_COMPUTE_ERROR_ON(t == nullptr);
+ os << R"([label = ")" << t->desc().shape << R"( \n )" << t->desc().data_type << R"("])";
+ os << ";\n";
+ }
+ }
+}
+} // namespace graph2
+} // namespace arm_compute
diff --git a/src/runtime/CL/CLSubTensor.cpp b/src/runtime/CL/CLSubTensor.cpp
index 5f58024b0e..d0e7d760ff 100644
--- a/src/runtime/CL/CLSubTensor.cpp
+++ b/src/runtime/CL/CLSubTensor.cpp
@@ -29,6 +29,11 @@
using namespace arm_compute;
+CLSubTensor::CLSubTensor()
+ : _parent(nullptr), _info()
+{
+}
+
CLSubTensor::CLSubTensor(ICLTensor *parent, const TensorShape &tensor_shape, const Coordinates &coords, bool extend_parent)
: _parent(nullptr), _info()
{
diff --git a/src/runtime/CL/functions/CLGEMM.cpp b/src/runtime/CL/functions/CLGEMM.cpp
index e6f8f266d8..c1926a78d1 100644
--- a/src/runtime/CL/functions/CLGEMM.cpp
+++ b/src/runtime/CL/functions/CLGEMM.cpp
@@ -138,8 +138,10 @@ void CLGEMM::configure(const ICLTensor *a, const ICLTensor *b, const ICLTensor *
// Manage intermediate buffers
_memory_group.manage(&_tmp_a);
- _memory_group.manage(&_tmp_b);
-
+ if(_reshape_b_only_on_first_run)
+ {
+ _memory_group.manage(&_tmp_b);
+ }
// _tmp_a and _tmp_b will be auto configured in _interleave_kernel and in _transpose_kernel
// Configure interleave kernel
diff --git a/src/runtime/CL/functions/CLGEMMConvolutionLayer.cpp b/src/runtime/CL/functions/CLGEMMConvolutionLayer.cpp
index 3cba98ce0e..bc339f176f 100644
--- a/src/runtime/CL/functions/CLGEMMConvolutionLayer.cpp
+++ b/src/runtime/CL/functions/CLGEMMConvolutionLayer.cpp
@@ -38,8 +38,8 @@
using namespace arm_compute;
using namespace arm_compute::misc::shape_calculator;
-CLConvolutionLayerReshapeWeights::CLConvolutionLayerReshapeWeights(std::shared_ptr<IMemoryManager> memory_manager)
- : _memory_group(std::move(memory_manager)), _weights_reshape_kernel(), _weights_transposed_kernel(), _weights_reshaped()
+CLConvolutionLayerReshapeWeights::CLConvolutionLayerReshapeWeights()
+ : _weights_reshape_kernel()
{
}
@@ -86,16 +86,12 @@ Status CLConvolutionLayerReshapeWeights::validate(const ITensorInfo *weights, co
void CLConvolutionLayerReshapeWeights::run()
{
- _memory_group.acquire();
-
CLScheduler::get().enqueue(_weights_reshape_kernel);
-
- _memory_group.release();
}
CLGEMMConvolutionLayer::CLGEMMConvolutionLayer(std::shared_ptr<IMemoryManager> memory_manager)
: _memory_group(memory_manager), _reshape_weights(), _im2col_kernel(), _mm_gemm(memory_manager), _mm_gemmlowp(memory_manager), _gemmlowp_output_stage(), _col2im_kernel(), _im2col_output(),
- _interleave_output(), _weights_reshaped(), _weights_transposed(), _gemm_output(), _tmp_output(), _is_quantized(false), _is_first_run(true)
+ _weights_reshaped(), _gemm_output(), _tmp_output(), _is_quantized(false), _is_first_run(true)
{
}
diff --git a/src/runtime/NEON/functions/NEConvolutionLayer.cpp b/src/runtime/NEON/functions/NEConvolutionLayer.cpp
index 0a491589ff..c16ce9b960 100644
--- a/src/runtime/NEON/functions/NEConvolutionLayer.cpp
+++ b/src/runtime/NEON/functions/NEConvolutionLayer.cpp
@@ -30,6 +30,7 @@
#include <cmath>
#include <tuple>
+#include <utility>
namespace arm_compute
{
diff --git a/src/runtime/SubTensor.cpp b/src/runtime/SubTensor.cpp
index c5b8f33c9a..b010a32eca 100644
--- a/src/runtime/SubTensor.cpp
+++ b/src/runtime/SubTensor.cpp
@@ -27,6 +27,11 @@
using namespace arm_compute;
+SubTensor::SubTensor()
+ : _parent(nullptr), _info()
+{
+}
+
SubTensor::SubTensor(ITensor *parent, const TensorShape &tensor_shape, const Coordinates &coords, bool extend_parent)
: _parent(nullptr), _info()
{