Introduce Context opaque object of the new interface

An AclContext is introduced as part of the new interface.
This object is responsible for any constructural services that the
operators and other objects might need.

Main options that can be passed to a context object are:
- a target: for which all the subsequent object should bind with
- capabilities: which are the isa/target features to enable
- a mode: for which different strategies can be selected in the backend

Signed-off-by: Georgios Pinitas <georgios.pinitas@arm.com>
Change-Id: I315549e55d4d064cbe94dfa29d070dc281b447de
Reviewed-on: https://review.mlplatform.org/c/ml/ComputeLibrary/+/5088
Tested-by: Arm Jenkins <bsgcomp@arm.com>
Comments-Addressed: Arm Jenkins <bsgcomp@arm.com>

43 files changed, 1707 insertions, 27 deletions

diff --git a/src/c/AclContext.cpp b/src/c/AclContext.cpp
new file mode 100644
index 0000000000..e88995bcff
--- /dev/null
+++ b/src/c/AclContext.cpp
@@ -0,0 +1,122 @@
+/*
+ * Copyright (c) 2021 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/AclEntrypoints.h"
+
+#include "src/common/IContext.h"
+#include "src/common/utils/Macros.h"
+#include "src/common/utils/Validate.h"
+
+#ifdef ARM_COMPUTE_CPU_ENABLED
+#include "src/cpu/CpuContext.h"
+#endif /* ARM_COMPUTE_CPU_ENABLED */
+
+#ifdef ARM_COMPUTE_OPENCL_ENABLED
+#include "src/gpu/cl/ClContext.h"
+#endif /* ARM_COMPUTE_OPENCL_ENABLED */
+
+namespace
+{
+template <typename ContextType>
+arm_compute::IContext *create_backend_ctx(const AclContextOptions *options)
+{
+    return new(std::nothrow) ContextType(options);
+}
+
+bool is_target_valid(AclTarget target)
+{
+    return arm_compute::utils::is_in(target, { AclCpu, AclGpuOcl });
+}
+
+bool are_context_options_valid(const AclContextOptions *options)
+{
+    ARM_COMPUTE_ASSERT_NOT_NULLPTR(options);
+    return arm_compute::utils::is_in(options->mode, { AclPreferFastRerun, AclPreferFastStart });
+}
+
+arm_compute::IContext *create_context(AclTarget target, const AclContextOptions *options)
+{
+    switch(target)
+    {
+#ifdef ARM_COMPUTE_CPU_ENABLED
+        case AclCpu:
+            return create_backend_ctx<arm_compute::cpu::CpuContext>(options);
+#endif /* ARM_COMPUTE_CPU_ENABLED */
+#ifdef ARM_COMPUTE_OPENCL_ENABLED
+        case AclGpuOcl:
+            return create_backend_ctx<arm_compute::gpu::opencl::ClContext>(options);
+#endif /* ARM_COMPUTE_OPENCL_ENABLED */
+        default:
+            return nullptr;
+    }
+    return nullptr;
+}
+} // namespace
+
+extern "C" AclStatus AclCreateContext(AclContext              *ctx,
+                                      AclTarget                target,
+                                      const AclContextOptions *options)
+{
+    if(!is_target_valid(target))
+    {
+        ARM_COMPUTE_LOG_ERROR_ACL("Target is invalid");
+        return AclUnsupportedTarget;
+    }
+
+    if(options != nullptr && !are_context_options_valid(options))
+    {
+        ARM_COMPUTE_LOG_ERROR_ACL("Context options are invalid");
+        return AclInvalidArgument;
+    }
+
+    auto acl_ctx = create_context(target, options);
+    if(ctx == nullptr)
+    {
+        ARM_COMPUTE_LOG_ERROR_ACL("Couldn't allocate internal resources for context creation");
+        return AclOutOfMemory;
+    }
+    *ctx = acl_ctx;
+
+    return AclSuccess;
+}
+
+extern "C" AclStatus AclDestroyContext(AclContext external_ctx)
+{
+    using namespace arm_compute;
+
+    IContext *ctx = get_internal(external_ctx);
+
+    StatusCode status = StatusCode::Success;
+    status            = detail::validate_internal_context(ctx);
+    ARM_COMPUTE_RETURN_CENUM_ON_FAILURE(status);
+
+    if(ctx->refcount() != 0)
+    {
+        ARM_COMPUTE_LOG_ERROR_ACL("Context has references on it that haven't been released");
+        // TODO: Fix the refcount with callback when reaches 0
+    }
+
+    delete ctx;
+
+    return utils::as_cenum<AclStatus>(status);
+}
diff --git a/src/c/AclVersion.cpp b/src/c/AclVersion.cpp
new file mode 100644
index 0000000000..971189a6d4
--- /dev/null
+++ b/src/c/AclVersion.cpp
@@ -0,0 +1,42 @@
+/*
+ * Copyright (c) 2021 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/AclVersion.h"
+
+namespace
+{
+constexpr AclVersion version_info
+{
+    ARM_COMPUTE_LIBRARY_VERSION_MAJOR,
+    ARM_COMPUTE_LIBRARY_VERSION_MINOR,
+    ARM_COMPUTE_LIBRARY_VERSION_PATCH,
+#ifndef DOXYGEN_SKIP_THIS
+#include "arm_compute_version.embed"
+#endif /* DOXYGEN_SKIP_THIS */
+};
+} // namespace
+
+extern "C" const AclVersion *AclVersionInfo()
+{
+    return &version_info;
+}
diff --git a/src/c/cl/AclOpenClExt.cpp b/src/c/cl/AclOpenClExt.cpp
new file mode 100644
index 0000000000..5f2bb47c16
--- /dev/null
+++ b/src/c/cl/AclOpenClExt.cpp
@@ -0,0 +1,83 @@
+/*
+ * Copyright (c) 2021 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/AclOpenClExt.h"
+
+#include "src/common/Types.h"
+#include "src/gpu/cl/ClContext.h"
+
+#include "support/Cast.h"
+
+extern "C" AclStatus AclGetClContext(AclContext external_ctx, cl_context *opencl_context)
+{
+    using namespace arm_compute;
+    IContext *ctx = get_internal(external_ctx);
+
+    if(detail::validate_internal_context(ctx) != StatusCode::Success)
+    {
+        return AclStatus::AclInvalidArgument;
+    }
+
+    if(ctx->type() != Target::GpuOcl)
+    {
+        return AclStatus::AclInvalidTarget;
+    }
+
+    if(opencl_context == nullptr)
+    {
+        return AclStatus::AclInvalidArgument;
+    }
+
+    *opencl_context = utils::cast::polymorphic_downcast<arm_compute::gpu::opencl::ClContext *>(ctx)->cl_ctx().get();
+
+    return AclStatus::AclSuccess;
+}
+
+extern "C" AclStatus AclSetClContext(AclContext external_ctx, cl_context opencl_context)
+{
+    using namespace arm_compute;
+    IContext *ctx = get_internal(external_ctx);
+
+    if(detail::validate_internal_context(ctx) != StatusCode::Success)
+    {
+        return AclStatus::AclInvalidArgument;
+    }
+
+    if(ctx->type() != Target::GpuOcl)
+    {
+        return AclStatus::AclInvalidTarget;
+    }
+
+    if(ctx->refcount() != 0)
+    {
+        return AclStatus::AclUnsupportedConfig;
+    }
+
+    auto cl_ctx = utils::cast::polymorphic_downcast<arm_compute::gpu::opencl::ClContext *>(ctx);
+    if(!cl_ctx->set_cl_ctx(::cl::Context(opencl_context)))
+    {
+        return AclStatus::AclRuntimeError;
+    }
+
+    return AclStatus::AclSuccess;
+}
\ No newline at end of file
diff --git a/src/common/AllocatorWrapper.cpp b/src/common/AllocatorWrapper.cpp
new file mode 100644
index 0000000000..7b5bb34433
--- /dev/null
+++ b/src/common/AllocatorWrapper.cpp
@@ -0,0 +1,65 @@
+/*
+ * Copyright (c) 2021 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 "src/common/AllocatorWrapper.h"
+#include "arm_compute/core/Error.h"
+
+namespace arm_compute
+{
+AllocatorWrapper::AllocatorWrapper(const AclAllocator &backing_allocator) noexcept
+    : _backing_allocator(backing_allocator)
+{
+}
+
+void *AllocatorWrapper::alloc(size_t size)
+{
+    ARM_COMPUTE_ERROR_ON(_backing_allocator.alloc == nullptr);
+    return _backing_allocator.alloc(_backing_allocator.user_data, size);
+}
+
+void AllocatorWrapper::free(void *ptr)
+{
+    ARM_COMPUTE_ERROR_ON(_backing_allocator.free == nullptr);
+    _backing_allocator.free(_backing_allocator.user_data, ptr);
+}
+
+void *AllocatorWrapper::aligned_alloc(size_t size, size_t alignment)
+{
+    ARM_COMPUTE_ERROR_ON(_backing_allocator.aligned_alloc == nullptr);
+    return _backing_allocator.aligned_alloc(_backing_allocator.user_data, size, alignment);
+}
+
+void AllocatorWrapper::aligned_free(void *ptr)
+{
+    ARM_COMPUTE_ERROR_ON(_backing_allocator.aligned_free == nullptr);
+    _backing_allocator.aligned_free(_backing_allocator.user_data, ptr);
+}
+
+void AllocatorWrapper::set_user_data(void *user_data)
+{
+    if(user_data != nullptr)
+    {
+        _backing_allocator.user_data = user_data;
+    }
+}
+} // namespace arm_compute
diff --git a/src/common/AllocatorWrapper.h b/src/common/AllocatorWrapper.h
new file mode 100644
index 0000000000..5e1f138f16
--- /dev/null
+++ b/src/common/AllocatorWrapper.h
@@ -0,0 +1,81 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_ALLOCATORWRAPPER_H
+#define SRC_COMMON_ALLOCATORWRAPPER_H
+
+#include "arm_compute/AclTypes.h"
+
+namespace arm_compute
+{
+/** Default malloc allocator implementation */
+class AllocatorWrapper final
+{
+public:
+    /** Default Constructor
+     *
+     * @param[in] backing_allocator Backing memory allocator to be used
+     */
+    AllocatorWrapper(const AclAllocator &backing_allocator) noexcept;
+    AllocatorWrapper(const AllocatorWrapper &) noexcept = default;
+    AllocatorWrapper(AllocatorWrapper &&) noexcept      = default;
+    AllocatorWrapper &operator=(const AllocatorWrapper &) noexcept = delete;
+    AllocatorWrapper &operator=(AllocatorWrapper &&other) noexcept = default;
+    /** Allocate a chunk of memory of a given size in bytes
+     *
+     * @param[in] size Size of memory to allocate in bytes
+     *
+     * @return A pointer to the allocated memory if successful else nullptr
+     */
+    void *alloc(size_t size);
+    /** Free an allocated memory block
+     *
+     * @param[in] ptr Pointer to allocated memory
+     */
+    void free(void *ptr);
+    /** Allocate a chunk of memory of a given size in bytes,
+     *  while honoring a given alignment requirement
+     *
+     * @param[in] size      Size of memory to allocate in bytes
+     * @param[in] alignment Alignment requirements
+     *
+     * @return A pointer to the allocated memory if successful else nullptr
+     */
+    void *aligned_alloc(size_t size, size_t alignment);
+    /** Free an aligned memory block
+     *
+     * @param[in] ptr Pointer to the memory to release
+     */
+    void aligned_free(void *ptr);
+    /** Set user data to be used by the allocator
+     *
+     * @param[in] user_data User data to be used by the allocator
+     */
+    void set_user_data(void *user_data);
+
+private:
+    AclAllocator _backing_allocator;
+};
+} // namespace arm_compute
+
+#endif /* SRC_COMMON_ALLOCATORWRAPPER_H */
\ No newline at end of file
diff --git a/src/common/IContext.h b/src/common/IContext.h
new file mode 100644
index 0000000000..0d23abd2be
--- /dev/null
+++ b/src/common/IContext.h
@@ -0,0 +1,127 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_ICONTEXT_H
+#define SRC_COMMON_ICONTEXT_H
+
+#include "src/common/Types.h"
+#include "src/common/utils/Log.h"
+#include "src/common/utils/Object.h"
+
+#include <atomic>
+
+struct AclContext_
+{
+    arm_compute::detail::Header header{ arm_compute::detail::ObjectType::Context, nullptr };
+
+protected:
+    AclContext_()  = default;
+    ~AclContext_() = default;
+};
+
+namespace arm_compute
+{
+/**< Context interface */
+class IContext : public AclContext_
+{
+public:
+    IContext(Target target)
+        : AclContext_(), _target(target), _refcount(0)
+    {
+    }
+    /** Virtual Destructor */
+    virtual ~IContext()
+    {
+        header.type = detail::ObjectType::Invalid;
+    };
+    /** Target type accessor
+     *
+     * @return Target that the context is associated with
+     */
+    Target type() const
+    {
+        return _target;
+    }
+    /** Increment context refcount */
+    void inc_ref() const
+    {
+        ++_refcount;
+    }
+    /** Decrement context refcount */
+    void dec_ref() const
+    {
+        --_refcount;
+    }
+    /** Reference counter accessor
+     *
+     * @return The number of references pointing to this object
+     */
+    int refcount() const
+    {
+        return _refcount;
+    }
+    /** Checks if an object is valid
+     *
+     * @return True if sucessful otherwise false
+     */
+    bool is_valid() const
+    {
+        return header.type == detail::ObjectType::Context;
+    }
+
+private:
+    Target                   _target;   /**< Target type of context */
+    mutable std::atomic<int> _refcount; /**< Reference counter */
+};
+
+/** Extract internal representation of a Context
+ *
+ * @param[in] ctx Opaque context pointer
+ *
+ * @return The internal representation as an IContext
+ */
+inline IContext *get_internal(AclContext ctx)
+{
+    return static_cast<IContext *>(ctx);
+}
+
+namespace detail
+{
+/** Check if an internal context is valid
+ *
+ * @param[in] ctx Internal context to check
+ *
+ * @return A status code
+ */
+inline StatusCode validate_internal_context(const IContext *ctx)
+{
+    if(ctx == nullptr || !ctx->is_valid())
+    {
+        ARM_COMPUTE_LOG_ERROR_ACL("Invalid context object");
+        return StatusCode::InvalidArgument;
+    }
+    return StatusCode::Success;
+}
+} // namespace detail
+} // namespace arm_compute
+#endif /* SRC_COMMON_ICONTEXT_H */
diff --git a/src/common/Types.h b/src/common/Types.h
new file mode 100644
index 0000000000..60a11b04ec
--- /dev/null
+++ b/src/common/Types.h
@@ -0,0 +1,56 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_TYPES_H_
+#define SRC_COMMON_TYPES_H_
+
+#include "arm_compute/AclTypes.h"
+
+namespace arm_compute
+{
+enum class StatusCode
+{
+    Success            = AclSuccess,
+    RuntimeError       = AclRuntimeError,
+    OutOfMemory        = AclOutOfMemory,
+    Unimplemented      = AclUnimplemented,
+    UnsupportedTarget  = AclUnsupportedTarget,
+    InvalidTarget      = AclInvalidTarget,
+    InvalidArgument    = AclInvalidArgument,
+    UnsupportedConfig  = AclUnsupportedConfig,
+    InvalidObjectState = AclInvalidObjectState,
+};
+
+enum class Target
+{
+    Cpu    = AclTarget::AclCpu,
+    GpuOcl = AclTarget::AclGpuOcl,
+};
+
+enum class ExecutionMode
+{
+    FastRerun = AclPreferFastRerun,
+    FastStart = AclPreferFastStart,
+};
+} // namespace arm_compute
+#endif /* SRC_COMMON_TYPES_H_ */
diff --git a/src/common/utils/Log.h b/src/common/utils/Log.h
new file mode 100644
index 0000000000..0d6a50da92
--- /dev/null
+++ b/src/common/utils/Log.h
@@ -0,0 +1,80 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_LOG_H
+#define SRC_COMMON_LOG_H
+
+#include "arm_compute/core/Error.h"
+#include "arm_compute/core/utils/logging/Macros.h"
+
+#ifdef ARM_COMPUTE_LOGGING_ENABLED
+/** Create a logger
+ *
+ * @note It will eventually create all default loggers in don't exist
+ */
+#define ARM_COMPUTE_CREATE_ACL_LOGGER()                                                                                        \
+    do                                                                                                                         \
+    {                                                                                                                          \
+        if(arm_compute::logging::LoggerRegistry::get().logger("ComputeLibrary") == nullptr)                                    \
+        {                                                                                                                      \
+            arm_compute::logging::LoggerRegistry::get().create_logger("ComputeLibrary", arm_compute::logging::LogLevel::INFO); \
+        }                                                                                                                      \
+    } while(false)
+#else /* ARM_COMPUTE_LOGGING_ENABLED */
+#define ARM_COMPUTE_CREATE_ACL_LOGGER()
+#endif /* ARM_COMPUTE_LOGGING_ENABLED */
+/** Log a message to the logger
+ *
+ * @param[in] log_level Logging level
+ * @param[in] msg       Message to log
+ */
+#define ARM_COMPUTE_LOG_MSG_ACL(log_level, msg)                \
+    do                                                         \
+    {                                                          \
+        ARM_COMPUTE_CREATE_ACL_LOGGER();                       \
+        ARM_COMPUTE_LOG_MSG("ComputeLibrary", log_level, msg); \
+    } while(false)
+/** Log a message with format to the logger
+ *
+ * @param[in] log_level Logging level
+ * @param[in] fmt       String format (printf style)
+ * @param[in] ...       Message arguments
+ */
+#define ARM_COMPUTE_LOG_MSG_WITH_FORMAT_ACL(log_level, fmt, ...)                        \
+    do                                                                                  \
+    {                                                                                   \
+        ARM_COMPUTE_CREATE_ACL_LOGGER();                                                \
+        ARM_COMPUTE_LOG_MSG_WITH_FORMAT("ComputeLibrary", log_level, fmt, __VA_ARGS__); \
+    } while(false)
+/** Log an error message to the logger
+ *
+ * @param[in] msg Message to log
+ */
+#define ARM_COMPUTE_LOG_ERROR_ACL(msg)                                                     \
+    do                                                                                     \
+    {                                                                                      \
+        ARM_COMPUTE_CREATE_ACL_LOGGER();                                                   \
+        ARM_COMPUTE_LOG_MSG("ComputeLibrary", arm_compute::logging::LogLevel::ERROR, msg); \
+    } while(false)
+
+#endif /* SRC_COMMON_LOG_H */
diff --git a/src/common/utils/Macros.h b/src/common/utils/Macros.h
new file mode 100644
index 0000000000..2e44ea599e
--- /dev/null
+++ b/src/common/utils/Macros.h
@@ -0,0 +1,37 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_MACROS_H_
+#define SRC_COMMON_MACROS_H_
+
+#include "src/common/utils/Utils.h"
+
+#define ARM_COMPUTE_RETURN_CENUM_ON_FAILURE(status)                 \
+    {                                                               \
+        if(status != arm_compute::StatusCode::Success)              \
+        {                                                           \
+            return arm_compute::utils::as_cenum<AclStatus>(status); \
+        }                                                           \
+    }
+
+#endif /* SRC_COMMON_MACROS_H_ */
diff --git a/src/common/utils/Object.h b/src/common/utils/Object.h
new file mode 100644
index 0000000000..6f56f77d3c
--- /dev/null
+++ b/src/common/utils/Object.h
@@ -0,0 +1,65 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_OBJECT_H_
+#define SRC_COMMON_OBJECT_H_
+
+#include <cstdint>
+
+namespace arm_compute
+{
+// Forward declarations
+class IContext;
+
+namespace detail
+{
+/**< Object type enumerations */
+enum class ObjectType : uint32_t
+{
+    Context    = 1,
+    Queue      = 2,
+    Tensor     = 3,
+    TensorPack = 4,
+    Operator   = 5,
+    Invalid    = 0x56DEAD78
+};
+
+/**< API Header meta-data construct used by all opaque constructs */
+struct Header
+{
+    /** Constructor
+     *
+     * @param[in] type_ Object identification type
+     * @param[in] ctx_  Context to reference
+     */
+    Header(ObjectType type_, IContext *ctx_)
+        : type(type_), ctx(ctx_)
+    {
+    }
+
+    ObjectType type{ ObjectType::Invalid };
+    IContext *ctx{ nullptr };
+};
+} // namespace detail
+} // namespace arm_compute
+#endif /* SRC_COMMON_OBJECT_H_ */
diff --git a/src/common/utils/Utils.h b/src/common/utils/Utils.h
new file mode 100644
index 0000000000..9602c32f62
--- /dev/null
+++ b/src/common/utils/Utils.h
@@ -0,0 +1,67 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_COMMON_UTILS_H
+#define SRC_COMMON_UTILS_H
+
+#include <type_traits>
+
+namespace arm_compute
+{
+namespace utils
+{
+/** Convert a strongly typed enum to an old plain c enum
+ *
+ * @tparam E  Plain old C enum
+ * @tparam SE Strongly typed resulting enum
+ *
+ * @param[in] v Value to convert
+ *
+ * @return A corresponding plain old C enumeration
+ */
+template <typename E, typename SE>
+constexpr E as_cenum(SE v) noexcept
+{
+    return static_cast<E>(static_cast<std::underlying_type_t<SE>>(v));
+}
+/** Check if the given value is in the given enum value list
+ *
+ * @tparam E  The type of the enum
+ *
+ * @param[in] check Value to check
+ * @param[in] list  List of enum values to check against
+ *
+ * @return True if the given value is found in the list
+ */
+template <typename E>
+bool is_in(E check, std::initializer_list<E> list)
+{
+    return std::any_of(std::cbegin(list), std::cend(list), [&check](E e)
+    {
+        return check == e;
+    });
+}
+} // namespace utils
+} // namespace arm_compute
+
+#endif /* SRC_COMMON_UTILS_H */
diff --git a/src/core/common/Validate.h b/src/common/utils/Validate.h
index fa24bf5fa7..4e8807273a 100644
--- a/src/core/common/Validate.h
+++ b/src/common/utils/Validate.h
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020 Arm Limited.
+ * Copyright (c) 2020-2021 Arm Limited.
  *
  * SPDX-License-Identifier: MIT
  *
diff --git a/src/core/NEON/kernels/NELogicalKernel.cpp b/src/core/NEON/kernels/NELogicalKernel.cpp
index 27605e15c6..d98694ffe1 100644
--- a/src/core/NEON/kernels/NELogicalKernel.cpp
+++ b/src/core/NEON/kernels/NELogicalKernel.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2020 Arm Limited.
+ * Copyright (c) 2020-2021 Arm Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -25,7 +25,7 @@
 
 #include "arm_compute/core/Helpers.h"
 #include "arm_compute/core/Validate.h"
-#include "src/core/common/Validate.h"
+#include "src/common/utils/Validate.h"
 #include "src/core/helpers/AutoConfiguration.h"
 #include "src/core/helpers/WindowHelpers.h"
 
diff --git a/src/core/NEON/kernels/batchnormalization/impl/NEON/fp16.cpp b/src/core/NEON/kernels/batchnormalization/impl/NEON/fp16.cpp
index 4108d79a5a..c105adac70 100644
--- a/src/core/NEON/kernels/batchnormalization/impl/NEON/fp16.cpp
+++ b/src/core/NEON/kernels/batchnormalization/impl/NEON/fp16.cpp
@@ -27,7 +27,6 @@
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/kernels/detail/NEActivationFunctionDetail.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/NEON/kernels/batchnormalization/impl/NEON/fp32.cpp b/src/core/NEON/kernels/batchnormalization/impl/NEON/fp32.cpp
index 1fdc5bd5b1..4a90a211c7 100644
--- a/src/core/NEON/kernels/batchnormalization/impl/NEON/fp32.cpp
+++ b/src/core/NEON/kernels/batchnormalization/impl/NEON/fp32.cpp
@@ -27,7 +27,6 @@
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/kernels/detail/NEActivationFunctionDetail.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/NEON/kernels/batchnormalization/impl/SVE/fp16.cpp b/src/core/NEON/kernels/batchnormalization/impl/SVE/fp16.cpp
index 5638dcef0a..3e3e81d044 100644
--- a/src/core/NEON/kernels/batchnormalization/impl/SVE/fp16.cpp
+++ b/src/core/NEON/kernels/batchnormalization/impl/SVE/fp16.cpp
@@ -25,7 +25,6 @@
 #include "arm_compute/core/ITensorPack.h"
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/SVEMath.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/NEON/kernels/batchnormalization/impl/SVE/fp32.cpp b/src/core/NEON/kernels/batchnormalization/impl/SVE/fp32.cpp
index 51397aca3a..b0d4cbb684 100644
--- a/src/core/NEON/kernels/batchnormalization/impl/SVE/fp32.cpp
+++ b/src/core/NEON/kernels/batchnormalization/impl/SVE/fp32.cpp
@@ -25,7 +25,6 @@
 #include "arm_compute/core/ITensorPack.h"
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/SVEMath.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/NEON/kernels/scale/impl/NEON/fp16.cpp b/src/core/NEON/kernels/scale/impl/NEON/fp16.cpp
new file mode 100644
index 0000000000..0ad66cab1c
--- /dev/null
+++ b/src/core/NEON/kernels/scale/impl/NEON/fp16.cpp
@@ -0,0 +1,174 @@
+/*
+ * Copyright (c) 2021 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/core/Helpers.h"
+#include "arm_compute/core/ITensorPack.h"
+#include "arm_compute/core/Window.h"
+#include "src/core/NEON/NEMath.h"
+#include "src/core/NEON/wrapper/wrapper.h"
+#include "src/core/helpers/ScaleHelpers.h"
+#include "src/core/utils/ScaleUtils.h"
+#include "support/Rounding.h"
+
+#include <arm_neon.h>
+#include <cmath>
+#include <cstddef>
+
+#if defined(__ARM_FEATURE_FP16_VECTOR_ARITHMETIC) && defined(ENABLE_FP16_KERNELS)
+
+namespace arm_compute
+{
+namespace
+{
+void fp16_neon_scale_nearest(const ITensor *src, ITensor *dst, const ITensor *offsets,
+                             float sampling_offset, bool align_corners, const Window &window)
+{
+    const size_t in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const size_t in_stride_w  = src->info()->dimension(1) + src->info()->padding().top + src->info()->padding().bottom;
+    const size_t in_stride_wc = in_stride_w * in_stride_c;
+    const size_t in_dim_h     = src->info()->dimension(2);
+
+    // Compute the ratio between source height and destination height
+    const auto hr             = scale_utils::calculate_resize_ratio(in_dim_h, dst->info()->dimension(2), align_corners);
+    const auto window_start_x = static_cast<int32_t>(window.x().start());
+    const auto window_end_x   = static_cast<int32_t>(window.x().end());
+    const int  window_step_x  = 8;
+
+    Window win(window);
+    win.set(Window::DimX, Window::Dimension(0, 1, 1));
+    Iterator out(dst, win);
+
+    const uint8_t     *in_ptr_start        = src->buffer() + src->info()->offset_first_element_in_bytes();
+    const unsigned int in_stride_bytes_hwc = src->info()->strides_in_bytes()[3];
+
+    execute_window_loop(win, [&](const Coordinates & id)
+    {
+        const int32_t    offset     = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z()))) * in_stride_c;
+        const auto       in_hi      = static_cast<int>(align_corners ? utils::rounding::round_half_away_from_zero((id.z() + sampling_offset) * hr) : std::floor((id.z() + sampling_offset) * hr));
+        const int        offset_row = in_hi * in_stride_wc;
+        int32_t          x          = window_start_x;
+        const float16_t *in_ptr     = reinterpret_cast<const float16_t *>(in_ptr_start + in_stride_bytes_hwc * id[3]);
+
+        for(; x <= window_end_x - window_step_x; x += window_step_x)
+        {
+            wrapper::vstore(reinterpret_cast<float16_t *>(out.ptr()) + x,
+                            wrapper::vloadq(in_ptr + offset + offset_row + x));
+        }
+        for(; x < window_end_x; ++x)
+        {
+            *(reinterpret_cast<float16_t *>(out.ptr()) + x) = *(in_ptr + offset + offset_row + x);
+        }
+    },
+    out);
+}
+
+void fp16_neon_scale_bilinear(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                              BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                              bool align_corners, const Window &window)
+{
+    // Compute the ratio between source height and destination height
+    const auto hr = scale_utils::calculate_resize_ratio(src->info()->dimension(2), dst->info()->dimension(2), align_corners);
+
+    Iterator  out(dst, window);
+    const int in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const int in_dim_w     = src->info()->dimension(1);
+    const int in_dim_h     = src->info()->dimension(2);
+    const int in_stride_wc = in_stride_c * (in_dim_w + src->info()->padding().top + src->info()->padding().bottom);
+
+    // Don't increment in Y and Z direction for the input tensor
+    // A pointer to the start of this plane is needed as base for the precomputed offsets
+    Window win_in(window);
+    win_in.set(Window::DimY, Window::Dimension(0, 0, 0));
+    win_in.set(Window::DimZ, Window::Dimension(0, 0, 0));
+    Iterator in(src, win_in);
+
+    if(border_mode == BorderMode::CONSTANT)
+    {
+        using ConstType = typename std::conditional<std::is_same<float16_t, float16_t>::value, half, float16_t>::type;
+
+        const float16_t const_border_value = static_cast<float16_t>(constant_border_value.get<ConstType>());
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto       offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto       dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto       dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int32_t    in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+            const float16_t *in_ptr = reinterpret_cast<const float16_t *>(in.ptr()) + offset * in_stride_c + in_hi * in_stride_wc;
+
+            const auto a00 = (0 <= offset && offset < in_dim_w && 0 <= in_hi && in_hi < in_dim_h) ? *in_ptr : const_border_value;
+            const auto a01 = (-1 <= offset && offset < in_dim_w - 1 && 0 <= in_hi && in_hi < in_dim_h) ? *(in_ptr + in_stride_c) : const_border_value;
+            const auto a10 = (0 <= offset && offset < in_dim_w && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_wc) : const_border_value;
+            const auto a11 = (-1 <= offset && offset < in_dim_w - 1 && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_c + in_stride_wc) : const_border_value;
+
+            *reinterpret_cast<float16_t *>(out.ptr()) = static_cast<float16_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else if(border_mode == BorderMode::REPLICATE)
+    {
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int  in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+
+            auto clamped_w  = utility::clamp<int>(offset, 0, in_dim_w - 1);
+            auto clamped_w1 = utility::clamp<int>(offset + 1, 0, in_dim_w - 1);
+            auto clamped_h  = utility::clamp<int>(in_hi, 0, in_dim_h - 1);
+            auto clamped_h1 = utility::clamp<int>(in_hi + 1, 0, in_dim_h - 1);
+
+            const auto a00 = *(reinterpret_cast<const float16_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h * in_stride_wc);
+            const auto a01 = *(reinterpret_cast<const float16_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h * in_stride_wc);
+            const auto a10 = *(reinterpret_cast<const float16_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h1 * in_stride_wc);
+            const auto a11 = *(reinterpret_cast<const float16_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h1 * in_stride_wc);
+
+            *reinterpret_cast<float16_t *>(out.ptr()) = static_cast<float16_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else
+    {
+        ARM_COMPUTE_ERROR("Not implemented");
+    }
+}
+}
+namespace cpu
+{
+void fp16_neon_scale(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                     InterpolationPolicy policy, BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                     bool align_corners, const Window &window)
+{
+    if(policy == InterpolationPolicy::BILINEAR)
+    {
+        fp16_neon_scale_bilinear(src, dst, offsets, dx, dy, border_mode, constant_border_value, sampling_offset, align_corners, window);
+    }
+    else if(policy == InterpolationPolicy::NEAREST_NEIGHBOR)
+    {
+        fp16_neon_scale_nearest(src, dst, offsets, sampling_offset, align_corners, window);
+    }
+}
+} // namespace cpu
+} // namespace arm_compute
+
+#endif /* defined(__ARM_FEATURE_FP16_VECTOR_ARITHMETIC) && defined(ENABLE_FP16_KERNELS) */
\ No newline at end of file
diff --git a/src/core/NEON/kernels/scale/impl/NEON/integer.cpp b/src/core/NEON/kernels/scale/impl/NEON/integer.cpp
new file mode 100644
index 0000000000..a2359aac94
--- /dev/null
+++ b/src/core/NEON/kernels/scale/impl/NEON/integer.cpp
@@ -0,0 +1,293 @@
+/*
+ * Copyright (c) 2021 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/core/Helpers.h"
+#include "arm_compute/core/ITensorPack.h"
+#include "arm_compute/core/Window.h"
+#include "src/core/NEON/NEMath.h"
+#include "src/core/NEON/wrapper/wrapper.h"
+#include "src/core/helpers/ScaleHelpers.h"
+#include "src/core/utils/ScaleUtils.h"
+#include "support/Rounding.h"
+
+#include <arm_neon.h>
+#include <cmath>
+#include <cstddef>
+
+namespace arm_compute
+{
+namespace
+{
+void u8_neon_scale_nearest(const ITensor *src, ITensor *dst, const ITensor *offsets,
+                           float sampling_offset, bool align_corners, const Window &window)
+{
+    const size_t in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const size_t in_stride_w  = src->info()->dimension(1) + src->info()->padding().top + src->info()->padding().bottom;
+    const size_t in_stride_wc = in_stride_w * in_stride_c;
+    const size_t in_dim_h     = src->info()->dimension(2);
+
+    // Compute the ratio between source height and destination height
+    const auto hr             = scale_utils::calculate_resize_ratio(in_dim_h, dst->info()->dimension(2), align_corners);
+    const auto window_start_x = static_cast<int32_t>(window.x().start());
+    const auto window_end_x   = static_cast<int32_t>(window.x().end());
+    const int  window_step_x  = 16;
+
+    Window win(window);
+    win.set(Window::DimX, Window::Dimension(0, 1, 1));
+    Iterator out(dst, win);
+
+    const uint8_t     *in_ptr_start        = src->buffer() + src->info()->offset_first_element_in_bytes();
+    const unsigned int in_stride_bytes_hwc = src->info()->strides_in_bytes()[3];
+
+    execute_window_loop(win, [&](const Coordinates & id)
+    {
+        const int32_t  offset     = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z()))) * in_stride_c;
+        const auto     in_hi      = static_cast<int>(align_corners ? utils::rounding::round_half_away_from_zero((id.z() + sampling_offset) * hr) : std::floor((id.z() + sampling_offset) * hr));
+        const int      offset_row = in_hi * in_stride_wc;
+        int32_t        x          = window_start_x;
+        const uint8_t *in_ptr     = reinterpret_cast<const uint8_t *>(in_ptr_start + in_stride_bytes_hwc * id[3]);
+
+        for(; x <= window_end_x - window_step_x; x += window_step_x)
+        {
+            wrapper::vstore(reinterpret_cast<uint8_t *>(out.ptr()) + x,
+                            wrapper::vloadq(in_ptr + offset + offset_row + x));
+        }
+        for(; x < window_end_x; ++x)
+        {
+            *(reinterpret_cast<uint8_t *>(out.ptr()) + x) = *(in_ptr + offset + offset_row + x);
+        }
+    },
+    out);
+}
+
+void u8_neon_scale_bilinear(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                            BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                            bool align_corners, const Window &window)
+{
+    // Compute the ratio between source height and destination height
+    const auto hr = scale_utils::calculate_resize_ratio(src->info()->dimension(2), dst->info()->dimension(2), align_corners);
+
+    Iterator  out(dst, window);
+    const int in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const int in_dim_w     = src->info()->dimension(1);
+    const int in_dim_h     = src->info()->dimension(2);
+    const int in_stride_wc = in_stride_c * (in_dim_w + src->info()->padding().top + src->info()->padding().bottom);
+
+    // Don't increment in Y and Z direction for the input tensor
+    // A pointer to the start of this plane is needed as base for the precomputed offsets
+    Window win_in(window);
+    win_in.set(Window::DimY, Window::Dimension(0, 0, 0));
+    win_in.set(Window::DimZ, Window::Dimension(0, 0, 0));
+    Iterator in(src, win_in);
+
+    if(border_mode == BorderMode::CONSTANT)
+    {
+        const uint8_t const_border_value = static_cast<uint8_t>(constant_border_value.get<uint8_t>());
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto     offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto     dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto     dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int32_t  in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+            const uint8_t *in_ptr = reinterpret_cast<const uint8_t *>(in.ptr()) + offset * in_stride_c + in_hi * in_stride_wc;
+
+            const auto a00 = (0 <= offset && offset < in_dim_w && 0 <= in_hi && in_hi < in_dim_h) ? *in_ptr : const_border_value;
+            const auto a01 = (-1 <= offset && offset < in_dim_w - 1 && 0 <= in_hi && in_hi < in_dim_h) ? *(in_ptr + in_stride_c) : const_border_value;
+            const auto a10 = (0 <= offset && offset < in_dim_w && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_wc) : const_border_value;
+            const auto a11 = (-1 <= offset && offset < in_dim_w - 1 && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_c + in_stride_wc) : const_border_value;
+
+            *reinterpret_cast<uint8_t *>(out.ptr()) = static_cast<uint8_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else if(border_mode == BorderMode::REPLICATE)
+    {
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int  in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+
+            auto clamped_w  = utility::clamp<int>(offset, 0, in_dim_w - 1);
+            auto clamped_w1 = utility::clamp<int>(offset + 1, 0, in_dim_w - 1);
+            auto clamped_h  = utility::clamp<int>(in_hi, 0, in_dim_h - 1);
+            auto clamped_h1 = utility::clamp<int>(in_hi + 1, 0, in_dim_h - 1);
+
+            const auto a00 = *(reinterpret_cast<const uint8_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h * in_stride_wc);
+            const auto a01 = *(reinterpret_cast<const uint8_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h * in_stride_wc);
+            const auto a10 = *(reinterpret_cast<const uint8_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h1 * in_stride_wc);
+            const auto a11 = *(reinterpret_cast<const uint8_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h1 * in_stride_wc);
+
+            *reinterpret_cast<uint8_t *>(out.ptr()) = static_cast<uint8_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else
+    {
+        ARM_COMPUTE_ERROR("Not implemented");
+    }
+}
+
+void s16_neon_scale_nearest(const ITensor *src, ITensor *dst, const ITensor *offsets,
+                            float sampling_offset, bool align_corners, const Window &window)
+{
+    const size_t in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const size_t in_stride_w  = src->info()->dimension(1) + src->info()->padding().top + src->info()->padding().bottom;
+    const size_t in_stride_wc = in_stride_w * in_stride_c;
+    const size_t in_dim_h     = src->info()->dimension(2);
+
+    // Compute the ratio between source height and destination height
+    const auto hr             = scale_utils::calculate_resize_ratio(in_dim_h, dst->info()->dimension(2), align_corners);
+    const auto window_start_x = static_cast<int32_t>(window.x().start());
+    const auto window_end_x   = static_cast<int32_t>(window.x().end());
+    const int  window_step_x  = 8;
+
+    Window win(window);
+    win.set(Window::DimX, Window::Dimension(0, 1, 1));
+    Iterator out(dst, win);
+
+    const uint8_t     *in_ptr_start        = src->buffer() + src->info()->offset_first_element_in_bytes();
+    const unsigned int in_stride_bytes_hwc = src->info()->strides_in_bytes()[3];
+
+    execute_window_loop(win, [&](const Coordinates & id)
+    {
+        const int32_t  offset     = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z()))) * in_stride_c;
+        const auto     in_hi      = static_cast<int>(align_corners ? utils::rounding::round_half_away_from_zero((id.z() + sampling_offset) * hr) : std::floor((id.z() + sampling_offset) * hr));
+        const int      offset_row = in_hi * in_stride_wc;
+        int32_t        x          = window_start_x;
+        const int16_t *in_ptr     = reinterpret_cast<const int16_t *>(in_ptr_start + in_stride_bytes_hwc * id[3]);
+
+        for(; x <= window_end_x - window_step_x; x += window_step_x)
+        {
+            wrapper::vstore(reinterpret_cast<int16_t *>(out.ptr()) + x,
+                            wrapper::vloadq(in_ptr + offset + offset_row + x));
+        }
+        for(; x < window_end_x; ++x)
+        {
+            *(reinterpret_cast<int16_t *>(out.ptr()) + x) = *(in_ptr + offset + offset_row + x);
+        }
+    },
+    out);
+}
+
+void s16_neon_scale_bilinear(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                             BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                             bool align_corners, const Window &window)
+{
+    // Compute the ratio between source height and destination height
+    const auto hr = scale_utils::calculate_resize_ratio(src->info()->dimension(2), dst->info()->dimension(2), align_corners);
+
+    Iterator  out(dst, window);
+    const int in_stride_c  = src->info()->dimension(0) + src->info()->padding().left + src->info()->padding().right;
+    const int in_dim_w     = src->info()->dimension(1);
+    const int in_dim_h     = src->info()->dimension(2);
+    const int in_stride_wc = in_stride_c * (in_dim_w + src->info()->padding().top + src->info()->padding().bottom);
+
+    // Don't increment in Y and Z direction for the input tensor
+    // A pointer to the start of this plane is needed as base for the precomputed offsets
+    Window win_in(window);
+    win_in.set(Window::DimY, Window::Dimension(0, 0, 0));
+    win_in.set(Window::DimZ, Window::Dimension(0, 0, 0));
+    Iterator in(src, win_in);
+
+    if(border_mode == BorderMode::CONSTANT)
+    {
+        const int16_t const_border_value = static_cast<int16_t>(constant_border_value.get<int16_t>());
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto     offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto     dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto     dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int32_t  in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+            const int16_t *in_ptr = reinterpret_cast<const int16_t *>(in.ptr()) + offset * in_stride_c + in_hi * in_stride_wc;
+
+            const auto a00 = (0 <= offset && offset < in_dim_w && 0 <= in_hi && in_hi < in_dim_h) ? *in_ptr : const_border_value;
+            const auto a01 = (-1 <= offset && offset < in_dim_w - 1 && 0 <= in_hi && in_hi < in_dim_h) ? *(in_ptr + in_stride_c) : const_border_value;
+            const auto a10 = (0 <= offset && offset < in_dim_w && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_wc) : const_border_value;
+            const auto a11 = (-1 <= offset && offset < in_dim_w - 1 && -1 <= in_hi && in_hi < in_dim_h - 1) ? *(in_ptr + in_stride_c + in_stride_wc) : const_border_value;
+
+            *reinterpret_cast<int16_t *>(out.ptr()) = static_cast<int16_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else if(border_mode == BorderMode::REPLICATE)
+    {
+        execute_window_loop(window, [&](const Coordinates & id)
+        {
+            const auto offset = *reinterpret_cast<const int32_t *>(offsets->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dx_val = *reinterpret_cast<const float *>(dx->ptr_to_element(Coordinates(id.y(), id.z())));
+            const auto dy_val = *reinterpret_cast<const float *>(dy->ptr_to_element(Coordinates(id.y(), id.z())));
+            const int  in_hi  = std::floor((id.z() + sampling_offset) * hr - sampling_offset);
+
+            auto clamped_w  = utility::clamp<int>(offset, 0, in_dim_w - 1);
+            auto clamped_w1 = utility::clamp<int>(offset + 1, 0, in_dim_w - 1);
+            auto clamped_h  = utility::clamp<int>(in_hi, 0, in_dim_h - 1);
+            auto clamped_h1 = utility::clamp<int>(in_hi + 1, 0, in_dim_h - 1);
+
+            const auto a00 = *(reinterpret_cast<const int16_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h * in_stride_wc);
+            const auto a01 = *(reinterpret_cast<const int16_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h * in_stride_wc);
+            const auto a10 = *(reinterpret_cast<const int16_t *>(in.ptr()) + clamped_w * in_stride_c + clamped_h1 * in_stride_wc);
+            const auto a11 = *(reinterpret_cast<const int16_t *>(in.ptr()) + clamped_w1 * in_stride_c + clamped_h1 * in_stride_wc);
+
+            *reinterpret_cast<int16_t *>(out.ptr()) = static_cast<int16_t>(scale_helpers::delta_bilinear(a00, a01, a10, a11, dx_val, dy_val));
+        },
+        in, out);
+    }
+    else
+    {
+        ARM_COMPUTE_ERROR("Not implemented");
+    }
+}
+}
+namespace cpu
+{
+void u8_neon_scale(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                   InterpolationPolicy policy, BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                   bool align_corners, const Window &window)
+{
+    if(policy == InterpolationPolicy::BILINEAR)
+    {
+        u8_neon_scale_bilinear(src, dst, offsets, dx, dy, border_mode, constant_border_value, sampling_offset, align_corners, window);
+    }
+    else if(policy == InterpolationPolicy::NEAREST_NEIGHBOR)
+    {
+        u8_neon_scale_nearest(src, dst, offsets, sampling_offset, align_corners, window);
+    }
+}
+
+void s16_neon_scale(const ITensor *src, ITensor *dst, const ITensor *offsets, const ITensor *dx, const ITensor *dy,
+                    InterpolationPolicy policy, BorderMode border_mode, PixelValue constant_border_value, float sampling_offset,
+                    bool align_corners, const Window &window)
+{
+    if(policy == InterpolationPolicy::BILINEAR)
+    {
+        s16_neon_scale_bilinear(src, dst, offsets, dx, dy, border_mode, constant_border_value, sampling_offset, align_corners, window);
+    }
+    else if(policy == InterpolationPolicy::NEAREST_NEIGHBOR)
+    {
+        s16_neon_scale_nearest(src, dst, offsets, sampling_offset, align_corners, window);
+    }
+}
+} // namespace cpu
+} // namespace arm_compute
\ No newline at end of file
diff --git a/src/core/NEON/kernels/scale/impl/NEON/list.h b/src/core/NEON/kernels/scale/impl/NEON/list.h
index 11cac26071..c91242f5b2 100644
--- a/src/core/NEON/kernels/scale/impl/NEON/list.h
+++ b/src/core/NEON/kernels/scale/impl/NEON/list.h
@@ -29,7 +29,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
 #include "support/Rounding.h"
diff --git a/src/core/NEON/kernels/scale/impl/NEON/qasymm8_signed.cpp b/src/core/NEON/kernels/scale/impl/NEON/qasymm8_signed.cpp
index 7b1860802f..149cdf478f 100644
--- a/src/core/NEON/kernels/scale/impl/NEON/qasymm8_signed.cpp
+++ b/src/core/NEON/kernels/scale/impl/NEON/qasymm8_signed.cpp
@@ -21,7 +21,6 @@
  * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
  * SOFTWARE.
  */
-
 #include "src/core/NEON/kernels/scale/impl/NEON/list.h"
 
 namespace arm_compute
diff --git a/src/core/NEON/kernels/scale/impl/SVE/fp16.cpp b/src/core/NEON/kernels/scale/impl/SVE/fp16.cpp
index 91c3dc3b4e..99f08dbdf9 100644
--- a/src/core/NEON/kernels/scale/impl/SVE/fp16.cpp
+++ b/src/core/NEON/kernels/scale/impl/SVE/fp16.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
 #include "support/Rounding.h"
diff --git a/src/core/NEON/kernels/scale/impl/SVE/fp32.cpp b/src/core/NEON/kernels/scale/impl/SVE/fp32.cpp
index abb4faa6c5..94055ae953 100644
--- a/src/core/NEON/kernels/scale/impl/SVE/fp32.cpp
+++ b/src/core/NEON/kernels/scale/impl/SVE/fp32.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
 #include "support/Rounding.h"
diff --git a/src/core/NEON/kernels/scale/impl/SVE/integer.cpp b/src/core/NEON/kernels/scale/impl/SVE/integer.cpp
index 5f5263c5d3..2a724ece31 100644
--- a/src/core/NEON/kernels/scale/impl/SVE/integer.cpp
+++ b/src/core/NEON/kernels/scale/impl/SVE/integer.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
 #include "support/Rounding.h"
diff --git a/src/core/NEON/kernels/scale/impl/SVE/qasymm8.cpp b/src/core/NEON/kernels/scale/impl/SVE/qasymm8.cpp
index fc65ff44f0..c475ad615c 100644
--- a/src/core/NEON/kernels/scale/impl/SVE/qasymm8.cpp
+++ b/src/core/NEON/kernels/scale/impl/SVE/qasymm8.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
diff --git a/src/core/NEON/kernels/scale/impl/SVE/qasymm8_signed.cpp b/src/core/NEON/kernels/scale/impl/SVE/qasymm8_signed.cpp
index 676ca94fb0..b39b75abba 100644
--- a/src/core/NEON/kernels/scale/impl/SVE/qasymm8_signed.cpp
+++ b/src/core/NEON/kernels/scale/impl/SVE/qasymm8_signed.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/helpers/ScaleHelpers.h"
 #include "src/core/utils/ScaleUtils.h"
diff --git a/src/core/cpu/kernels/activation/NEON/fp16.cpp b/src/core/cpu/kernels/activation/NEON/fp16.cpp
index 7fe4ab3f63..0ddd43ea0e 100644
--- a/src/core/cpu/kernels/activation/NEON/fp16.cpp
+++ b/src/core/cpu/kernels/activation/NEON/fp16.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Helpers.h"
 #include "arm_compute/core/Validate.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/activation/NEON/fp32.cpp b/src/core/cpu/kernels/activation/NEON/fp32.cpp
index f1f2753813..244ca5739f 100644
--- a/src/core/cpu/kernels/activation/NEON/fp32.cpp
+++ b/src/core/cpu/kernels/activation/NEON/fp32.cpp
@@ -26,7 +26,6 @@
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/activation/NEON/qasymm8.cpp b/src/core/cpu/kernels/activation/NEON/qasymm8.cpp
index 7506a8294f..a1217435b6 100644
--- a/src/core/cpu/kernels/activation/NEON/qasymm8.cpp
+++ b/src/core/cpu/kernels/activation/NEON/qasymm8.cpp
@@ -27,7 +27,6 @@
 #include "src/core/NEON/NEAsymm.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/activation/NEON/qasymm8_signed.cpp b/src/core/cpu/kernels/activation/NEON/qasymm8_signed.cpp
index 8f75abea8e..8b40bf8e72 100644
--- a/src/core/cpu/kernels/activation/NEON/qasymm8_signed.cpp
+++ b/src/core/cpu/kernels/activation/NEON/qasymm8_signed.cpp
@@ -26,7 +26,6 @@
 #include "src/core/NEON/NEAsymm.h"
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/activation/NEON/qsymm16.cpp b/src/core/cpu/kernels/activation/NEON/qsymm16.cpp
index 9eee360427..54b41820f2 100644
--- a/src/core/cpu/kernels/activation/NEON/qsymm16.cpp
+++ b/src/core/cpu/kernels/activation/NEON/qsymm16.cpp
@@ -28,7 +28,6 @@
 #include "src/core/NEON/NEMath.h"
 #include "src/core/NEON/NESymm.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/activation/SVE/fp16.cpp b/src/core/cpu/kernels/activation/SVE/fp16.cpp
index 8208813cd3..bf31fd7d93 100644
--- a/src/core/cpu/kernels/activation/SVE/fp16.cpp
+++ b/src/core/cpu/kernels/activation/SVE/fp16.cpp
@@ -24,7 +24,6 @@
 #include "arm_compute/core/Helpers.h"
 #include "arm_compute/core/ITensorPack.h"
 #include "arm_compute/core/Window.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/cpu/kernels/activation/SVE/fp32.cpp b/src/core/cpu/kernels/activation/SVE/fp32.cpp
index 55bdc9999e..75f9f8a4c3 100644
--- a/src/core/cpu/kernels/activation/SVE/fp32.cpp
+++ b/src/core/cpu/kernels/activation/SVE/fp32.cpp
@@ -25,7 +25,6 @@
 #include "arm_compute/core/ITensorPack.h"
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/SVEMath.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/cpu/kernels/activation/SVE/qasymm8.cpp b/src/core/cpu/kernels/activation/SVE/qasymm8.cpp
index 3d9476ac56..228b4ae530 100644
--- a/src/core/cpu/kernels/activation/SVE/qasymm8.cpp
+++ b/src/core/cpu/kernels/activation/SVE/qasymm8.cpp
@@ -24,7 +24,6 @@
 
 #include "arm_compute/core/Helpers.h"
 #include "arm_compute/core/Window.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/cpu/kernels/activation/SVE/qasymm8_signed.cpp b/src/core/cpu/kernels/activation/SVE/qasymm8_signed.cpp
index 0b3d798942..989f825eb9 100644
--- a/src/core/cpu/kernels/activation/SVE/qasymm8_signed.cpp
+++ b/src/core/cpu/kernels/activation/SVE/qasymm8_signed.cpp
@@ -24,7 +24,6 @@
 #include "arm_compute/core/Helpers.h"
 #include "arm_compute/core/Window.h"
 #include "src/core/NEON/wrapper/wrapper.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/cpu/kernels/activation/SVE/qsymm16.cpp b/src/core/cpu/kernels/activation/SVE/qsymm16.cpp
index dbaf267bf9..66974875da 100644
--- a/src/core/cpu/kernels/activation/SVE/qsymm16.cpp
+++ b/src/core/cpu/kernels/activation/SVE/qsymm16.cpp
@@ -25,7 +25,6 @@
 #include "arm_compute/core/ITensorPack.h"
 #include "arm_compute/core/Window.h"
 #include "arm_compute/core/experimental/Types.h"
-#include "src/core/common/Validate.h"
 
 #include <cmath>
 #include <cstddef>
diff --git a/src/core/cpu/kernels/floor/NEON/fp16.cpp b/src/core/cpu/kernels/floor/NEON/fp16.cpp
index 0d31eb77f8..f362676a36 100644
--- a/src/core/cpu/kernels/floor/NEON/fp16.cpp
+++ b/src/core/cpu/kernels/floor/NEON/fp16.cpp
@@ -23,8 +23,8 @@
  */
 #if defined(__ARM_FEATURE_FP16_VECTOR_ARITHMETIC) && defined(ENABLE_FP16_KERNELS)
 
+#include "src/common/utils/Validate.h"
 #include "src/core/NEON/NEMath.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/cpu/kernels/floor/NEON/fp32.cpp b/src/core/cpu/kernels/floor/NEON/fp32.cpp
index dd63f9f9d7..f5efb2e849 100644
--- a/src/core/cpu/kernels/floor/NEON/fp32.cpp
+++ b/src/core/cpu/kernels/floor/NEON/fp32.cpp
@@ -21,8 +21,8 @@
  * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
  * SOFTWARE.
  */
+#include "src/common/utils/Validate.h"
 #include "src/core/NEON/NEMath.h"
-#include "src/core/common/Validate.h"
 
 #include <arm_neon.h>
 #include <cmath>
diff --git a/src/core/gpu/cl/kernels/ClElementwiseKernel.cpp b/src/core/gpu/cl/kernels/ClElementwiseKernel.cpp
index 7d204b1348..8f12eb2215 100644
--- a/src/core/gpu/cl/kernels/ClElementwiseKernel.cpp
+++ b/src/core/gpu/cl/kernels/ClElementwiseKernel.cpp
@@ -25,8 +25,8 @@
 
 #include "arm_compute/core/CL/CLHelpers.h"
 #include "arm_compute/core/CL/ICLTensor.h"
+#include "src/common/utils/Validate.h"
 #include "src/core/CL/CLValidate.h"
-#include "src/core/common/Validate.h"
 #include "src/core/helpers/AutoConfiguration.h"
 #include "src/core/helpers/WindowHelpers.h"
 #include "support/Cast.h"
diff --git a/src/cpu/CpuContext.cpp b/src/cpu/CpuContext.cpp
new file mode 100644
index 0000000000..5980a5ab0d
--- /dev/null
+++ b/src/cpu/CpuContext.cpp
@@ -0,0 +1,182 @@
+/*
+ * Copyright (c) 2021 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 "src/cpu/CpuContext.h"
+
+#include "arm_compute/core/CPP/CPPTypes.h"
+#include "src/runtime/CPUUtils.h"
+
+#include <cstdlib>
+
+namespace arm_compute
+{
+namespace cpu
+{
+namespace
+{
+void *default_allocate(void *user_data, size_t size)
+{
+    ARM_COMPUTE_UNUSED(user_data);
+    return ::operator new(size);
+}
+void default_free(void *user_data, void *ptr)
+{
+    ARM_COMPUTE_UNUSED(user_data);
+    ::operator delete(ptr);
+}
+void *default_aligned_allocate(void *user_data, size_t size, size_t alignment)
+{
+    ARM_COMPUTE_UNUSED(user_data);
+    void *ptr = nullptr;
+#if defined(BARE_METAL) || defined(__APPLE__)
+    size_t rem       = size % alignment;
+    size_t real_size = (rem) ? (size + alignment - rem) : size;
+    ptr              = aligned_alloc(alignment, real_size);
+#else  /* defined(BARE_METAL) || defined(__APPLE__) */
+    posix_memalign(&ptr, alignment, size);
+#endif /* defined(BARE_METAL) || defined(__APPLE__) */
+    return ptr;
+}
+void default_aligned_free(void *user_data, void *ptr)
+{
+    ARM_COMPUTE_UNUSED(user_data);
+    free(ptr);
+}
+static AclAllocator default_allocator = { &default_allocate,
+                                          &default_free,
+                                          &default_aligned_allocate,
+                                          &default_aligned_free,
+                                          nullptr
+                                        };
+
+AllocatorWrapper populate_allocator(AclAllocator *external_allocator)
+{
+    bool is_valid = (external_allocator != nullptr);
+    if(is_valid)
+    {
+        is_valid = is_valid && (external_allocator->alloc != nullptr);
+        is_valid = is_valid && (external_allocator->free != nullptr);
+        is_valid = is_valid && (external_allocator->aligned_alloc != nullptr);
+        is_valid = is_valid && (external_allocator->aligned_free != nullptr);
+    }
+    return is_valid ? AllocatorWrapper(*external_allocator) : AllocatorWrapper(default_allocator);
+}
+
+CpuCapabilities populate_capabilities_legacy(const CPUInfo &cpu_info)
+{
+    CpuCapabilities caps;
+
+    // Extract SIMD extension
+    caps.neon = true;
+#ifdef SVE2
+    caps.sve2 = true;
+#endif /* SVE2 */
+    // Extract data-type support
+    caps.fp16 = cpu_info.has_fp16();
+#ifdef V8P6_BF
+    caps.bf16 = true;
+#endif /* V8P6_BF */
+
+    // Extract ISA extensions
+    caps.dot = cpu_info.has_dotprod();
+#ifdef MMLA_FP32
+    caps.mmla_fp = true;
+#endif /* MMLA_FP32 */
+#ifdef MMLA_INT8
+    caps.mmla_int8 = true;
+#endif /* MMLA_INT8 */
+
+    return caps;
+}
+
+CpuCapabilities populate_capabilities_flags(AclTargetCapabilities external_caps)
+{
+    CpuCapabilities caps;
+
+    // Extract SIMD extension
+    caps.neon = external_caps & AclCpuCapabilitiesNeon;
+    caps.sve  = external_caps & AclCpuCapabilitiesSve;
+    caps.sve2 = external_caps & AclCpuCapabilitiesSve2;
+    // Extract data-type support
+    caps.fp16 = external_caps & AclCpuCapabilitiesFp16;
+    caps.bf16 = external_caps & AclCpuCapabilitiesBf16;
+    // Extract ISA extensions
+    caps.dot       = external_caps & AclCpuCapabilitiesDot;
+    caps.mmla_fp   = external_caps & AclCpuCapabilitiesMmlaFp;
+    caps.mmla_int8 = external_caps & AclCpuCapabilitiesMmlaInt8;
+
+    return caps;
+}
+
+CpuCapabilities populate_capabilities(AclTargetCapabilities external_caps,
+                                      int32_t               max_threads)
+{
+    // Extract legacy structure
+    CPUInfo cpu_info;
+    arm_compute::utils::cpu::get_cpu_configuration(cpu_info);
+
+    CpuCapabilities caps;
+    if(external_caps != AclCpuCapabilitiesAuto)
+    {
+        caps = populate_capabilities_flags(external_caps);
+    }
+    else
+    {
+        caps = populate_capabilities_legacy(cpu_info);
+    }
+
+    // Set max number of threads
+#if defined(BARE_METAL)
+    ARM_COMPUTE_UNUSED(max_threads);
+    caps.max_threads = 1;
+#else  /* defined(BARE_METAL) */
+    caps.max_threads = (max_threads > 0) ? max_threads : std::thread::hardware_concurrency();
+#endif /* defined(BARE_METAL) */
+
+    return caps;
+}
+} // namespace
+
+CpuContext::CpuContext(const AclContextOptions *options)
+    : IContext(Target::Cpu),
+      _allocator(default_allocator),
+      _caps(populate_capabilities(AclCpuCapabilitiesAuto, -1))
+{
+    if(options != nullptr)
+    {
+        _allocator = populate_allocator(options->allocator);
+        _caps      = populate_capabilities(options->capabilities, options->max_compute_units);
+    }
+}
+
+const CpuCapabilities &CpuContext::capabilities() const
+{
+    return _caps;
+}
+
+AllocatorWrapper &CpuContext::allocator()
+{
+    return _allocator;
+}
+} // namespace cpu
+} // namespace arm_compute
diff --git a/src/cpu/CpuContext.h b/src/cpu/CpuContext.h
new file mode 100644
index 0000000000..81bab97b8e
--- /dev/null
+++ b/src/cpu/CpuContext.h
@@ -0,0 +1,77 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_CPU_CPUCONTEXT_H
+#define SRC_CPU_CPUCONTEXT_H
+
+#include "src/common/AllocatorWrapper.h"
+#include "src/common/IContext.h"
+
+namespace arm_compute
+{
+namespace cpu
+{
+/** Structure that encodes the CPU capabilities to be used */
+struct CpuCapabilities
+{
+    bool neon{ false };
+    bool sve{ false };
+    bool sve2{ false };
+
+    bool fp16{ false };
+    bool bf16{ false };
+    bool dot{ false };
+    bool mmla_int8{ false };
+    bool mmla_fp{ false };
+
+    int32_t max_threads{ -1 };
+};
+
+/** CPU context implementation class */
+class CpuContext final : public IContext
+{
+public:
+    /** Default Constructor
+     *
+     * @param[in] options Creational options
+     */
+    explicit CpuContext(const AclContextOptions *options);
+    /** Cpu Capabilities accessor
+     *
+     * @return The ISA capabilities to be used by the CPU
+     */
+    const CpuCapabilities &capabilities() const;
+    /** Backing memory allocator accessor
+     *
+     * @return Allocator that allocates CPU memory
+     */
+    AllocatorWrapper &allocator();
+
+private:
+    AllocatorWrapper _allocator;
+    CpuCapabilities  _caps;
+};
+} // namespace cpu
+} // namespace arm_compute
+
+#endif /* SRC_CPU_CPUCONTEXT_H */
\ No newline at end of file
diff --git a/src/gpu/cl/ClContext.cpp b/src/gpu/cl/ClContext.cpp
new file mode 100644
index 0000000000..2bd8b8dd0e
--- /dev/null
+++ b/src/gpu/cl/ClContext.cpp
@@ -0,0 +1,74 @@
+/*
+ * Copyright (c) 2021 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 "src/gpu/cl/ClContext.h"
+
+namespace arm_compute
+{
+namespace gpu
+{
+namespace opencl
+{
+namespace
+{
+mlgo::MLGOHeuristics populate_mlgo(const char *filename)
+{
+    mlgo::MLGOHeuristics heuristics;
+    bool                 status = heuristics.reload_from_file(filename);
+    return status ? std::move(heuristics) : mlgo::MLGOHeuristics();
+}
+} // namespace
+
+ClContext::ClContext(const AclContextOptions *options)
+    : IContext(Target::GpuOcl),
+      _mlgo_heuristics(),
+      _cl_context()
+{
+    if(options != nullptr)
+    {
+        _mlgo_heuristics = populate_mlgo(options->kernel_config_file);
+    }
+}
+
+const mlgo::MLGOHeuristics &ClContext::mlgo() const
+{
+    return _mlgo_heuristics;
+}
+
+::cl::Context ClContext::cl_ctx()
+{
+    return _cl_context;
+}
+
+bool ClContext::set_cl_ctx(::cl::Context ctx)
+{
+    if(this->refcount() == 0)
+    {
+        _cl_context = ctx;
+        return true;
+    }
+    return false;
+}
+} // namespace opencl
+} // namespace gpu
+} // namespace arm_compute
diff --git a/src/gpu/cl/ClContext.h b/src/gpu/cl/ClContext.h
new file mode 100644
index 0000000000..e3f16b1c3f
--- /dev/null
+++ b/src/gpu/cl/ClContext.h
@@ -0,0 +1,76 @@
+/*
+ * Copyright (c) 2021 Arm Limited.
+ *
+ * SPDX-License-Identifier: MIT
+ *
+ * Permission is hereby granted, free of charge, to any person obtaining a copy
+ * of this software and associated documentation files (the "Software"), to
+ * deal in the Software without restriction, including without limitation the
+ * rights to use, copy, modify, merge, publish, distribute, sublicense, and/or
+ * sell copies of the Software, and to permit persons to whom the Software is
+ * furnished to do so, subject to the following conditions:
+ *
+ * The above copyright notice and this permission notice shall be included in all
+ * copies or substantial portions of the Software.
+ *
+ * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
+ * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
+ * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
+ * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
+ * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
+ * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE
+ * SOFTWARE.
+ */
+#ifndef SRC_GPU_CL_CPUCONTEXT_H
+#define SRC_GPU_CL_CPUCONTEXT_H
+
+#include "src/common/IContext.h"
+#include "src/runtime/CL/mlgo/MLGOHeuristics.h"
+
+#include "arm_compute/core/CL/OpenCL.h"
+
+namespace arm_compute
+{
+namespace gpu
+{
+namespace opencl
+{
+/** OpenCL context implementation class */
+class ClContext final : public IContext
+{
+public:
+    /** Default Constructor
+     *
+     * @param[in] options Creational options
+     */
+    explicit ClContext(const AclContextOptions *options);
+    /** Extract MLGO heuristics
+     *
+     * @return Heuristics tree
+     */
+    const mlgo::MLGOHeuristics &mlgo() const;
+
+    /** Underlying cl context accessor
+     *
+     * @return the cl context used
+     */
+    ::cl::Context cl_ctx();
+    /** Update/inject an underlying cl context object
+     *
+     * @warning Context will be able to set if the object doesn't have any pending reference to other objects
+     *
+     * @param[in] ctx Underlying cl context to be used
+     *
+     * @return true if the context was set successfully else falseS
+     */
+    bool set_cl_ctx(::cl::Context ctx);
+
+private:
+    mlgo::MLGOHeuristics _mlgo_heuristics;
+    ::cl::Context        _cl_context;
+};
+} // namespace opencl
+} // namespace gpu
+} // namespace arm_compute
+
+#endif /* SRC_GPU_CL_CPUCONTEXT_H */
\ No newline at end of file