COMPMID-1534 - Fix GEMM and Magnitude test for FP16

On GEMM we had accuracy issue
On Magnitude we have disabled the fp16 acceleration since we do not have feature parity with CL
and this function is not used for ML

Change-Id: Iaebe3bbbd2a9f45db0c714aa5ebaf48eb0b65741
Reviewed-on: https://eu-gerrit-1.euhpc.arm.com/145467
Tested-by: Jenkins <bsgcomp@arm.com>
Reviewed-by: Georgios Pinitas <georgios.pinitas@arm.com>
Reviewed-by: Anthony Barbier <anthony.barbier@arm.com>

13 files changed, 56 insertions, 569 deletions

diff --git a/arm_compute/core/NEON/kernels/NEMagnitudePhaseKernel.h b/arm_compute/core/NEON/kernels/NEMagnitudePhaseKernel.h
index 696721673d..5a8355c793 100644
--- a/arm_compute/core/NEON/kernels/NEMagnitudePhaseKernel.h
+++ b/arm_compute/core/NEON/kernels/NEMagnitudePhaseKernel.h
@@ -97,77 +97,5 @@ private:
     ITensor                  *_magnitude; /**< Output - Magnitude */
     ITensor                  *_phase;     /**< Output - Phase */
 };
-
-#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-/** Template interface for the kernel to compute magnitude and phase */
-template <MagnitudeType mag_type, PhaseType phase_type>
-class NEMagnitudePhaseFP16Kernel : public INEKernel
-{
-public:
-    const char *name() const override
-    {
-        return "NEMagnitudePhaseFP16Kernel";
-    }
-    /** Default constructor */
-    NEMagnitudePhaseFP16Kernel();
-    /** Destructor */
-    ~NEMagnitudePhaseFP16Kernel() = default;
-    /** Prevent instances of this class from being copied (As this class contains pointers) */
-    NEMagnitudePhaseFP16Kernel(const NEMagnitudePhaseFP16Kernel &) = delete;
-    /** Default move constructor */
-    NEMagnitudePhaseFP16Kernel(NEMagnitudePhaseFP16Kernel &&) = default;
-    /** Prevent instances of this class from being copied (As this class contains pointers) */
-    NEMagnitudePhaseFP16Kernel &operator=(const NEMagnitudePhaseFP16Kernel &) = delete;
-    /** Default move assignment operator */
-    NEMagnitudePhaseFP16Kernel &operator=(NEMagnitudePhaseFP16Kernel &&) = default;
-
-    /** Initialise the kernel's input, output.
-     *
-     * @note At least one of out1 or out2 must be set
-     *
-     * @param[in]  gx        Gradient X tensor. Data type supported: S16.
-     * @param[in]  gy        Gradient Y tensor. Data type supported: S16.
-     * @param[out] magnitude (Optional) The output tensor - Magnitude. Data type supported: S16.
-     * @param[out] phase     (Optional) The output tensor - Phase. Data type supported: U8.
-     */
-    void configure(const ITensor *gx, const ITensor *gy, ITensor *magnitude, ITensor *phase);
-
-    // Inherited methods overridden:
-    void run(const Window &window, const ThreadInfo &info) override;
-
-private:
-    /** Function to perform magnitude on the given window
-     *
-     * @param[in] window Region on which to execute the kernel
-     */
-    void magnitude(const Window &window);
-    /** Function to perform phase on the given window
-     *
-     * @param[in] window Region on which to execute the kernel
-     */
-    void phase(const Window &window);
-    /** Function to perform magnitude and phase on the given window
-     *
-     * @param[in] window Region on which to execute the kernel
-     */
-    void magnitude_phase(const Window &window);
-
-    /** Common signature for all the specialised MagnitudePhase functions
-     *
-     * @param[in] window Region on which to execute the kernel.
-     */
-    using MagnitudePhaseFunctionPtr = void (NEMagnitudePhaseFP16Kernel::*)(const Window &window);
-    /** MagnitudePhase function to use for the particular formats passed to configure() */
-    MagnitudePhaseFunctionPtr _func;
-    const ITensor            *_gx;        /**< Input gradient X */
-    const ITensor            *_gy;        /**< Input gradient Y */
-    ITensor                  *_magnitude; /**< Output - Magnitude */
-    ITensor                  *_phase;     /**< Output - Phase */
-};
-#else  /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
-/** Template interface for the kernel to compute magnitude and phase */
-template <MagnitudeType mag_type, PhaseType phase_type>
-using NEMagnitudePhaseFP16Kernel = NEMagnitudePhaseKernel<mag_type, phase_type>;
-#endif /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
 } // namespace arm_compute
 #endif /* __ARM_COMPUTE_NEMAGNITUDEPHASEKERNEL_H__ */
diff --git a/arm_compute/runtime/CL/functions/CLMagnitude.h b/arm_compute/runtime/CL/functions/CLMagnitude.h
index f9c7e5c14a..9fd85f9a95 100644
--- a/arm_compute/runtime/CL/functions/CLMagnitude.h
+++ b/arm_compute/runtime/CL/functions/CLMagnitude.h
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -41,9 +41,8 @@ public:
      * @param[in]  input2   Second tensor input. Data types supported: S16.
      * @param[out] output   Output tensor. Data types supported: S16.
      * @param[in]  mag_type (Optional) Magnitude calculation type. Default: L2NORM.
-     * @param[in]  use_fp16 (Optional) If true the FP16 kernels will be used. If false F32 kernels are used.
      */
-    void configure(const ICLTensor *input1, const ICLTensor *input2, ICLTensor *output, MagnitudeType mag_type = MagnitudeType::L2NORM, bool use_fp16 = false);
+    void configure(const ICLTensor *input1, const ICLTensor *input2, ICLTensor *output, MagnitudeType mag_type = MagnitudeType::L2NORM);
 };
 }
 #endif /*__ARM_COMPUTE_CLMAGNITUDE_H__ */
diff --git a/arm_compute/runtime/NEON/functions/NEMagnitude.h b/arm_compute/runtime/NEON/functions/NEMagnitude.h
index 5bc3faf664..6aabe9dfa4 100644
--- a/arm_compute/runtime/NEON/functions/NEMagnitude.h
+++ b/arm_compute/runtime/NEON/functions/NEMagnitude.h
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -40,9 +40,8 @@ public:
      * @param[in]  input2   Second tensor input. Data type supported: S16.
      * @param[out] output   Output tensor. Data type supported: S16.
      * @param[in]  mag_type (Optional) Magnitude calculation type. Default: L2NORM.
-     * @param[in]  use_fp16 (Optional) If true the FP16 kernels will be used. If false F32 kernels are used.
      */
-    void configure(const ITensor *input1, const ITensor *input2, ITensor *output, MagnitudeType mag_type = MagnitudeType::L2NORM, bool use_fp16 = false);
+    void configure(const ITensor *input1, const ITensor *input2, ITensor *output, MagnitudeType mag_type = MagnitudeType::L2NORM);
 };
 }
 #endif /*__ARM_COMPUTE_NEMAGNITUDE_H__ */
diff --git a/src/core/NEON/kernels/NEMagnitudePhaseKernel.cpp b/src/core/NEON/kernels/NEMagnitudePhaseKernel.cpp
index 2d7c29d9a0..4a318f02c1 100644
--- a/src/core/NEON/kernels/NEMagnitudePhaseKernel.cpp
+++ b/src/core/NEON/kernels/NEMagnitudePhaseKernel.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -51,386 +51,6 @@ constexpr float COEFF1       = 0.0663f;
 constexpr float COEFF2       = 0.2447f;
 } // namespace
 
-#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-namespace fp16
-{
-inline float16x8_t inv(float16x8_t x)
-{
-    const float16x8_t estimate = vrecpeq_f16(x);
-    return vmulq_f16(estimate, vrecpsq_f16(x, estimate));
-}
-
-inline float16x8_t atan2_fast(float16x8_t gx, float16x8_t gy, float16x8_t scale)
-{
-    static const float16x8_t one     = vdupq_n_f16(1.0f);
-    static const float16x8_t ninety  = vdupq_n_f16(90.f * SCALE_FACTOR);
-    static const float16x8_t epsilon = vdupq_n_f16(1e-9f);
-    static const float16x8_t piover4 = vdupq_n_f16(PI_4);
-    static const float16x8_t coeff1  = vdupq_n_f16(COEFF1);
-    static const float16x8_t coeff2  = vdupq_n_f16(COEFF2);
-
-    const float16x8_t abs_gx = vabsq_f16(gx);
-    const float16x8_t abs_gy = vabsq_f16(gy);
-    const float16x8_t tmin   = vminq_f16(abs_gx, abs_gy);
-    const float16x8_t tmax   = vmaxq_f16(abs_gx, abs_gy);
-
-    // z = min(x, y) / max(x, y)
-    const float16x8_t z    = vmulq_f16(tmin, inv(vaddq_f16(tmax, epsilon)));
-    const float16x8_t absz = vabsq_f16(z);
-
-    //                   = x * [pi/4 + (1 - |x|) * (0.2447 + 0.0663 * |x|)]
-    float16x8_t arctan = vmulq_f16(z, vfmaq_f16(piover4,
-                                                vsubq_f16(one, absz),
-                                                vfmaq_f16(coeff2, coeff1, absz)));
-
-    // Radians to degrees conversion with applied a scale factor in order to have the result [0, 255]
-    arctan = vmulq_f16(arctan, scale);
-
-    /* If z > 1, result = 90 - result */
-    return vbslq_f16(vcgeq_f16(abs_gx, abs_gy), arctan, vsubq_f16(ninety, arctan));
-}
-
-inline float16x8_t atan2_0_360(float16x8_t gx, float16x8_t gy)
-{
-    static const float16x8_t scale      = vdupq_n_f16(SCALE_360);
-    static const float16x8_t threesixty = vdupq_n_f16(360.0f * SCALE_FACTOR);
-    static const float16x8_t zero       = vdupq_n_f16(0.0f);
-    static const float16x8_t oneeighty  = vdupq_n_f16(180.0f * SCALE_FACTOR);
-
-    float16x8_t arctan = atan2_fast(gx, gy, scale);
-
-    // Choose correct quadrant
-    arctan = vbslq_f16(vcltq_f16(gx, zero), vsubq_f16(oneeighty, arctan), arctan);
-    arctan = vbslq_f16(vcltq_f16(gy, zero), vsubq_f16(threesixty, arctan), arctan);
-
-    return arctan;
-}
-
-inline float16x8_t atan2_0_180(float16x8_t gx, float16x8_t gy)
-{
-    static const float16x8_t scale      = vdupq_n_f16(SCALE_180);
-    static const float16x8_t threesixty = vdupq_n_f16(360.0f * SCALE_FACTOR);
-    static const float16x8_t oneeighty  = vdupq_n_f16(180.0f * SCALE_FACTOR);
-    static const float16x8_t zero       = vdupq_n_f16(0.0f);
-
-    float16x8_t arctan = atan2_fast(gx, gy, scale);
-
-    // Choose correct quadrant
-    arctan = vbslq_f16(vcltq_f16(gx, zero), vsubq_f16(oneeighty, arctan), arctan);
-    arctan = vbslq_f16(vcltq_f16(gy, zero), vsubq_f16(threesixty, arctan), arctan);
-    arctan = vbslq_f16(vcgtq_f16(arctan, oneeighty), vsubq_f16(arctan, oneeighty), arctan);
-
-    return arctan;
-}
-
-inline float32x4_t invsqrtv(float32x4_t x)
-{
-    float32x4_t sqrt_reciprocal = vrsqrteq_f32(x);
-
-    sqrt_reciprocal = vmulq_f32(vrsqrtsq_f32(vmulq_f32(x, sqrt_reciprocal), sqrt_reciprocal),
-                                sqrt_reciprocal);
-    sqrt_reciprocal = vmulq_f32(vrsqrtsq_f32(vmulq_f32(x, sqrt_reciprocal), sqrt_reciprocal),
-                                sqrt_reciprocal);
-
-    return sqrt_reciprocal;
-}
-
-inline float32x4_t sqrtv(float32x4_t x)
-{
-    float32x4_t res = vdupq_n_f32(0.5f);
-    return vmlaq_f32(res, x, invsqrtv(x));
-}
-
-inline int16x8_t magnitude_l1(int16x8_t input1, int16x8_t input2)
-{
-    return vqaddq_s16(vqabsq_s16(input1), vqabsq_s16(input2));
-}
-
-inline int16x8_t magnitude_l2(int16x8_t input1, int16x8_t input2)
-{
-    const int32x4x2_t square_x =
-    {
-        vmull_s16(vget_low_s16(input1), vget_low_s16(input1)),
-        vmull_s16(vget_high_s16(input1), vget_high_s16(input1))
-    };
-
-    const int32x4x2_t square_y =
-    {
-        vmull_s16(vget_low_s16(input2), vget_low_s16(input2)),
-        vmull_s16(vget_high_s16(input2), vget_high_s16(input2))
-    };
-
-    const uint32x4x2_t sum =
-    {
-        vaddq_u32(vreinterpretq_u32_s32(square_x.val[0]),
-        vreinterpretq_u32_s32(square_y.val[0])),
-        vaddq_u32(vreinterpretq_u32_s32(square_x.val[1]),
-        vreinterpretq_u32_s32(square_y.val[1]))
-    };
-
-    const float32x4x2_t res =
-    {
-        sqrtv(vcvtq_f32_u32(sum.val[0])),
-        sqrtv(vcvtq_f32_u32(sum.val[1]))
-    };
-
-    return vcombine_s16(vqmovn_s32(vcvtq_s32_f32(res.val[0])),
-                        vqmovn_s32(vcvtq_s32_f32(res.val[1])));
-}
-
-inline uint8x8_t phase_signed(int16x8_t input1, int16x8_t input2)
-{
-    static const float16x8_t zeropointfive = vdupq_n_f16(0.5f);
-
-    const float16x8_t inputx_f16 = vcvtq_f16_s16(input1);
-    const float16x8_t inputy_f16 = vcvtq_f16_s16(input2);
-
-    // Compute fast atan2
-    const float16x8_t angle = atan2_0_360(inputx_f16, inputy_f16);
-
-    return vqmovun_s16(vcvtq_s16_f16(vaddq_f16(angle, zeropointfive)));
-}
-
-inline uint8x8_t phase_unsigned(int16x8_t input1, int16x8_t input2)
-{
-    static const float16x8_t zeropointfive = vdupq_n_f16(0.5f);
-
-    const float16x8_t inputx_f16 = vcvtq_f16_s16(input1);
-    const float16x8_t inputy_f16 = vcvtq_f16_s16(input2);
-
-    // Compute fast atan2
-    const float16x8_t angle = atan2_0_180(inputx_f16, inputy_f16);
-
-    return vqmovun_s16(vcvtq_s16_f16(vaddq_f16(angle, zeropointfive)));
-}
-
-template <MagnitudeType mag_type>
-inline int16x8x2_t compute_magnitude(const int16x8x2_t &in0, const int16x8x2_t &gx);
-
-template <>
-inline int16x8x2_t compute_magnitude<MagnitudeType::L2NORM>(const int16x8x2_t &in0, const int16x8x2_t &gx)
-{
-    const int16x8x2_t mag =
-    {
-        magnitude_l2(in0.val[0], gx.val[0]),
-        magnitude_l2(in0.val[1], gx.val[1])
-    };
-
-    return mag;
-}
-
-template <>
-inline int16x8x2_t compute_magnitude<MagnitudeType::L1NORM>(const int16x8x2_t &in0, const int16x8x2_t &gx)
-{
-    const int16x8x2_t mag =
-    {
-        magnitude_l1(in0.val[0], gx.val[0]),
-        magnitude_l1(in0.val[1], gx.val[1])
-    };
-
-    return mag;
-}
-
-template <PhaseType phase_type>
-inline uint8x16_t compute_phase(const int16x8x2_t &in0, const int16x8x2_t &gx);
-
-template <>
-inline uint8x16_t compute_phase<PhaseType::SIGNED>(const int16x8x2_t &in0, const int16x8x2_t &gx)
-{
-    return vcombine_u8(phase_signed(in0.val[0], gx.val[0]),
-                       phase_signed(in0.val[1], gx.val[1]));
-}
-
-template <>
-inline uint8x16_t compute_phase<PhaseType::UNSIGNED>(const int16x8x2_t &in0, const int16x8x2_t &gx)
-{
-    return vcombine_u8(phase_unsigned(in0.val[0], gx.val[0]),
-                       phase_unsigned(in0.val[1], gx.val[1]));
-}
-} // namespace fp16
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::NEMagnitudePhaseFP16Kernel()
-    : _func(nullptr), _gx(nullptr), _gy(nullptr), _magnitude(nullptr), _phase(nullptr)
-{
-}
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-void NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::configure(const ITensor *gx, const ITensor *gy, ITensor *magnitude, ITensor *phase)
-{
-    ARM_COMPUTE_ERROR_ON_FORMAT_NOT_IN(gx, Format::S16);
-    ARM_COMPUTE_ERROR_ON_FORMAT_NOT_IN(gy, Format::S16);
-    ARM_COMPUTE_ERROR_ON((nullptr == magnitude) && (nullptr == phase));
-
-    const bool run_mag   = magnitude != nullptr;
-    const bool run_phase = phase != nullptr;
-
-    if(run_mag)
-    {
-        ARM_COMPUTE_ERROR_ON_FORMAT_NOT_IN(magnitude, Format::S16);
-    }
-
-    if(run_phase)
-    {
-        ARM_COMPUTE_ERROR_ON_FORMAT_NOT_IN(phase, Format::U8);
-    }
-
-    _gx        = gx;
-    _gy        = gy;
-    _magnitude = magnitude;
-    _phase     = phase;
-
-    if(run_mag && run_phase)
-    {
-        /* Run magnitude and phase */
-        _func = &NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::magnitude_phase;
-    }
-    else if(run_mag)
-    {
-        /* Run magnitude */
-        _func = &NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::magnitude;
-    }
-    else if(run_phase)
-    {
-        /* Run phase */
-        _func = &NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::phase;
-    }
-    else
-    {
-        ARM_COMPUTE_ERROR("At least one output must be NOT NULL");
-    }
-
-    const unsigned int num_elems_processed_per_iteration = 16;
-
-    // Configure kernel window
-    Window                 win = calculate_max_window(*gx->info(), Steps(num_elems_processed_per_iteration));
-    AccessWindowHorizontal magnitude_access(magnitude == nullptr ? nullptr : magnitude->info(), 0, num_elems_processed_per_iteration);
-    AccessWindowHorizontal phase_access(phase == nullptr ? nullptr : phase->info(), 0, num_elems_processed_per_iteration);
-
-    update_window_and_padding(win,
-                              AccessWindowHorizontal(gx->info(), 0, num_elems_processed_per_iteration),
-                              AccessWindowHorizontal(gy->info(), 0, num_elems_processed_per_iteration),
-                              magnitude_access,
-                              phase_access);
-
-    ValidRegion valid_region = intersect_valid_regions(gx->info()->valid_region(),
-                                                       gy->info()->valid_region());
-
-    magnitude_access.set_valid_region(win, valid_region);
-    phase_access.set_valid_region(win, valid_region);
-
-    INEKernel::configure(win);
-}
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-void NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::magnitude(const Window &window)
-{
-    Iterator gx(_gx, window);
-    Iterator gy(_gy, window);
-    Iterator magnitude(_magnitude, window);
-
-    execute_window_loop(window, [&](const Coordinates & id)
-    {
-        const int16x8x2_t input1 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr()) + 8)
-        };
-
-        const int16x8x2_t input2 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr()) + 8)
-        };
-
-        // Compute and store magnitude
-        const int16x8x2_t mag = fp16::compute_magnitude<mag_type>(input1, input2);
-
-        /* Store magnitude */
-        vst1q_s16(reinterpret_cast<int16_t *>(magnitude.ptr()), mag.val[0]);
-        vst1q_s16(reinterpret_cast<int16_t *>(magnitude.ptr()) + 8, mag.val[1]);
-    },
-    gx, gy, magnitude);
-}
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-void NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::phase(const Window &window)
-{
-    Iterator gx(_gx, window);
-    Iterator gy(_gy, window);
-    Iterator phase(_phase, window);
-
-    execute_window_loop(window, [&](const Coordinates & id)
-    {
-        const int16x8x2_t input1 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr()) + 8)
-        };
-
-        const int16x8x2_t input2 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr()) + 8)
-        };
-
-        // Compute and store phase
-        vst1q_u8(phase.ptr(), fp16::compute_phase<phase_type>(input1, input2));
-    },
-    gx, gy, phase);
-}
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-void NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::magnitude_phase(const Window &window)
-{
-    Iterator gx(_gx, window);
-    Iterator gy(_gy, window);
-    Iterator magnitude(_magnitude, window);
-    Iterator phase(_phase, window);
-
-    execute_window_loop(window, [&](const Coordinates & id)
-    {
-        const int16x8x2_t input1 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gx.ptr()) + 8)
-        };
-
-        const int16x8x2_t input2 =
-        {
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr())),
-            vld1q_s16(reinterpret_cast<int16_t *>(gy.ptr()) + 8)
-        };
-
-        // Compute and store magnitude
-        const int16x8x2_t mag = fp16::compute_magnitude<mag_type>(input1, input2);
-
-        vst1q_s16(reinterpret_cast<int16_t *>(magnitude.ptr()), mag.val[0]);
-        vst1q_s16(reinterpret_cast<int16_t *>(magnitude.ptr()) + 8, mag.val[1]);
-
-        // Compute and store phase
-        vst1q_u8(phase.ptr(), fp16::compute_phase<phase_type>(input1, input2));
-    },
-    gx, gy, magnitude, phase);
-}
-
-template <MagnitudeType mag_type, PhaseType phase_type>
-void NEMagnitudePhaseFP16Kernel<mag_type, phase_type>::run(const Window &window, const ThreadInfo &info)
-{
-    ARM_COMPUTE_UNUSED(info);
-    ARM_COMPUTE_ERROR_ON_UNCONFIGURED_KERNEL(this);
-    ARM_COMPUTE_ERROR_ON_INVALID_SUBWINDOW(INEKernel::window(), window);
-    ARM_COMPUTE_ERROR_ON(_func == nullptr);
-
-    (this->*_func)(window);
-}
-
-template class arm_compute::NEMagnitudePhaseFP16Kernel<MagnitudeType::L1NORM, PhaseType::SIGNED>;
-template class arm_compute::NEMagnitudePhaseFP16Kernel<MagnitudeType::L2NORM, PhaseType::SIGNED>;
-template class arm_compute::NEMagnitudePhaseFP16Kernel<MagnitudeType::L1NORM, PhaseType::UNSIGNED>;
-template class arm_compute::NEMagnitudePhaseFP16Kernel<MagnitudeType::L2NORM, PhaseType::UNSIGNED>;
-#endif /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
-
 namespace
 {
 inline float32x4_t inv(float32x4_t x)
diff --git a/src/runtime/CL/functions/CLMagnitude.cpp b/src/runtime/CL/functions/CLMagnitude.cpp
index 9d6ac7a11a..e2dfe3a2a7 100644
--- a/src/runtime/CL/functions/CLMagnitude.cpp
+++ b/src/runtime/CL/functions/CLMagnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -30,10 +30,8 @@
 
 using namespace arm_compute;
 
-void CLMagnitude::configure(const ICLTensor *input1, const ICLTensor *input2, ICLTensor *output, MagnitudeType mag_type, bool use_fp16)
+void CLMagnitude::configure(const ICLTensor *input1, const ICLTensor *input2, ICLTensor *output, MagnitudeType mag_type)
 {
-    ARM_COMPUTE_UNUSED(use_fp16); //TODO(COMPMID-644): Add half float support
-
     auto k = arm_compute::support::cpp14::make_unique<CLMagnitudePhaseKernel>();
     k->configure(input1, input2, output, nullptr, mag_type);
     _kernel = std::move(k);
diff --git a/src/runtime/NEON/functions/NEMagnitude.cpp b/src/runtime/NEON/functions/NEMagnitude.cpp
index f86505449f..2738201d54 100644
--- a/src/runtime/NEON/functions/NEMagnitude.cpp
+++ b/src/runtime/NEON/functions/NEMagnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2016, 2017 ARM Limited.
+ * Copyright (c) 2016-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -31,36 +31,18 @@
 
 using namespace arm_compute;
 
-void NEMagnitude::configure(const ITensor *input1, const ITensor *input2, ITensor *output, MagnitudeType mag_type, bool use_fp16)
+void NEMagnitude::configure(const ITensor *input1, const ITensor *input2, ITensor *output, MagnitudeType mag_type)
 {
-    if(use_fp16)
+    if(mag_type == MagnitudeType::L1NORM)
     {
-        if(mag_type == MagnitudeType::L1NORM)
-        {
-            auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseFP16Kernel<MagnitudeType::L1NORM, PhaseType::SIGNED>>();
-            k->configure(input1, input2, output, nullptr);
-            _kernel = std::move(k);
-        }
-        else
-        {
-            auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseFP16Kernel<MagnitudeType::L2NORM, PhaseType::SIGNED>>();
-            k->configure(input1, input2, output, nullptr);
-            _kernel = std::move(k);
-        }
+        auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseKernel<MagnitudeType::L1NORM, PhaseType::SIGNED>>();
+        k->configure(input1, input2, output, nullptr);
+        _kernel = std::move(k);
     }
     else
     {
-        if(mag_type == MagnitudeType::L1NORM)
-        {
-            auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseKernel<MagnitudeType::L1NORM, PhaseType::SIGNED>>();
-            k->configure(input1, input2, output, nullptr);
-            _kernel = std::move(k);
-        }
-        else
-        {
-            auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseKernel<MagnitudeType::L2NORM, PhaseType::SIGNED>>();
-            k->configure(input1, input2, output, nullptr);
-            _kernel = std::move(k);
-        }
+        auto k = arm_compute::support::cpp14::make_unique<NEMagnitudePhaseKernel<MagnitudeType::L2NORM, PhaseType::SIGNED>>();
+        k->configure(input1, input2, output, nullptr);
+        _kernel = std::move(k);
     }
 }
diff --git a/tests/benchmark/CL/Magnitude.cpp b/tests/benchmark/CL/Magnitude.cpp
index 5e7508341a..c2bb590476 100644
--- a/tests/benchmark/CL/Magnitude.cpp
+++ b/tests/benchmark/CL/Magnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -47,12 +47,10 @@ using CLMagnitudeFixture = MagnitudeFixture<CLTensor, CLMagnitude, CLAccessor>;
 
 TEST_SUITE(CL)
 TEST_SUITE(Magnitude)
-REGISTER_FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::SmallImageShapes(), framework::dataset::make("Format", { Format::S16, Format::S32 })),
-                                                                                                                 magnitude_types),
-                                                                                                         framework::dataset::make("UseFP16", { false })));
-REGISTER_FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::LargeImageShapes(), framework::dataset::make("Format", { Format::S16, Format::S32 })),
-                                                                                                               magnitude_types),
-                                                                                                       framework::dataset::make("UseFP16", { false })));
+REGISTER_FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture, framework::DatasetMode::PRECOMMIT, combine(combine(datasets::SmallImageShapes(), framework::dataset::make("Format", { Format::S16, Format::S32 })),
+                                                                                                         magnitude_types));
+REGISTER_FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture, framework::DatasetMode::NIGHTLY, combine(combine(datasets::LargeImageShapes(), framework::dataset::make("Format", { Format::S16, Format::S32 })),
+                                                                                                       magnitude_types));
 TEST_SUITE_END() // Magnitude
 TEST_SUITE_END() // CL
 } // namespace benchmark
diff --git a/tests/benchmark/NEON/Magnitude.cpp b/tests/benchmark/NEON/Magnitude.cpp
index e2b1210453..d671b8e005 100644
--- a/tests/benchmark/NEON/Magnitude.cpp
+++ b/tests/benchmark/NEON/Magnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -47,25 +47,11 @@ using NEMagnitudeFixture = MagnitudeFixture<Tensor, NEMagnitude, Accessor>;
 
 TEST_SUITE(NEON)
 TEST_SUITE(Magnitude)
-
-#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-TEST_SUITE(FP16)
-REGISTER_FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::SmallImageShapes(), framework::dataset::make("Format", { Format::S16 })),
-                                                                                                                 magnitude_types),
-                                                                                                         framework::dataset::make("UseFP16", { true })));
-REGISTER_FIXTURE_DATA_TEST_CASE(RunLarge, NEMagnitudeFixture, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::LargeImageShapes(), framework::dataset::make("Format", { Format::S16 })),
-                                                                                                               magnitude_types),
-                                                                                                       framework::dataset::make("UseFP16", { true })));
-TEST_SUITE_END() // FP16
-#endif           // __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-
 TEST_SUITE(S16)
-REGISTER_FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::SmallImageShapes(), framework::dataset::make("Format", { Format::S16 })),
-                                                                                                                 magnitude_types),
-                                                                                                         framework::dataset::make("UseFP16", { false })));
-REGISTER_FIXTURE_DATA_TEST_CASE(RunLarge, NEMagnitudeFixture, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::LargeImageShapes(), framework::dataset::make("Format", { Format::S16 })),
-                                                                                                               magnitude_types),
-                                                                                                       framework::dataset::make("UseFP16", { false })));
+REGISTER_FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture, framework::DatasetMode::PRECOMMIT, combine(combine(datasets::SmallImageShapes(), framework::dataset::make("Format", { Format::S16 })),
+                                                                                                         magnitude_types));
+REGISTER_FIXTURE_DATA_TEST_CASE(RunLarge, NEMagnitudeFixture, framework::DatasetMode::NIGHTLY, combine(combine(datasets::LargeImageShapes(), framework::dataset::make("Format", { Format::S16 })),
+                                                                                                       magnitude_types));
 TEST_SUITE_END() // S16
 TEST_SUITE_END() // Magnitude
 TEST_SUITE_END() // NEON
diff --git a/tests/benchmark/fixtures/MagnitudeFixture.h b/tests/benchmark/fixtures/MagnitudeFixture.h
index f75540c66f..a1b8054529 100644
--- a/tests/benchmark/fixtures/MagnitudeFixture.h
+++ b/tests/benchmark/fixtures/MagnitudeFixture.h
@@ -41,7 +41,7 @@ class MagnitudeFixture : public framework::Fixture
 {
 public:
     template <typename...>
-    void setup(const TensorShape &shape, Format format, MagnitudeType magnitude_type, bool use_fp16)
+    void setup(const TensorShape &shape, Format format, MagnitudeType magnitude_type)
     {
         // Create tensors
         src1 = create_tensor<TensorType>(shape, format);
@@ -49,7 +49,7 @@ public:
         dst  = create_tensor<TensorType>(shape, format);
 
         // Create and configure function
-        magnitude_func.configure(&src1, &src2, &dst, magnitude_type, use_fp16);
+        magnitude_func.configure(&src1, &src2, &dst, magnitude_type);
 
         // Allocate tensors
         src1.allocator()->allocate();
diff --git a/tests/validation/CL/Magnitude.cpp b/tests/validation/CL/Magnitude.cpp
index b002239e01..7c517a44bb 100644
--- a/tests/validation/CL/Magnitude.cpp
+++ b/tests/validation/CL/Magnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -81,17 +81,15 @@ template <typename T>
 using CLMagnitudeFixture = MagnitudeValidationFixture<CLTensor, CLAccessor, CLMagnitude, T>;
 
 TEST_SUITE(S16)
-FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture<int16_t>, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S16)),
-                                                                                                                 framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                         framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture<int16_t>, framework::DatasetMode::PRECOMMIT, combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S16)),
+                                                                                                         framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
 {
     // Validate output
     validate(CLAccessor(_target), _reference, tolerance<int16_t>(_magnitude_type));
 }
 
-FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture<int16_t>, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S16)),
-                                                                                                               framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                       framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture<int16_t>, framework::DatasetMode::NIGHTLY, combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S16)),
+                                                                                                       framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
 {
     // Validate output
     validate(CLAccessor(_target), _reference, tolerance<int16_t>(_magnitude_type));
@@ -99,17 +97,15 @@ FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture<int16_t>, framework::Dataset
 TEST_SUITE_END() // S16
 
 TEST_SUITE(S32)
-FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture<int32_t>, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S32)),
-                                                                                                                 framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                         framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunSmall, CLMagnitudeFixture<int32_t>, framework::DatasetMode::PRECOMMIT, combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S32)),
+                                                                                                         framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
 {
     // Validate output
     validate(CLAccessor(_target), _reference, tolerance<int32_t>(_magnitude_type));
 }
 
-FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture<int32_t>, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S32)),
-                                                                                                               framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                       framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunLarge, CLMagnitudeFixture<int32_t>, framework::DatasetMode::NIGHTLY, combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S32)),
+                                                                                                       framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
 {
     // Validate output
     validate(CLAccessor(_target), _reference, tolerance<int32_t>(_magnitude_type));
diff --git a/tests/validation/NEON/GEMM.cpp b/tests/validation/NEON/GEMM.cpp
index 9c64131a61..ded5ec68c4 100644
--- a/tests/validation/NEON/GEMM.cpp
+++ b/tests/validation/NEON/GEMM.cpp
@@ -49,8 +49,12 @@ namespace validation
 {
 namespace
 {
-constexpr AbsoluteTolerance<float> tolerance_f(0.001f); /**< Tolerance value for comparing reference's output against implementation's output for floating point data types */
-
+constexpr AbsoluteTolerance<float> tolerance_f(0.001f); /**< Tolerance value for comparing reference's output against implementation's output for FP32 data types */
+#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
+RelativeTolerance<half_float::half> rel_tolerance_f16(half(0.2)); /**< Relative tolerance value for comparing reference's output against implementation's output for FP16 data types */
+const AbsoluteTolerance<float>      abs_tolerance_f16(0.2f);      /**< Absolute tolerance value for comparing reference's output against implementation's output for FP16 data types */
+constexpr float                     tolerance_num = 0.07f;        /**< Tolerance number for FP16 data types */
+#endif                                                            /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
 /** CNN data types */
 const auto CNNDataTypes = framework::dataset::make("DataType",
 {
@@ -125,13 +129,13 @@ TEST_SUITE(FP16)
 FIXTURE_DATA_TEST_CASE(RunSmall, NEGEMMFixture<half>, framework::DatasetMode::PRECOMMIT, combine(datasets::SmallGEMMDataset(), framework::dataset::make("DataType", DataType::F16)))
 {
     // Validate output
-    validate(Accessor(_target), _reference, tolerance_f);
+    validate(Accessor(_target), _reference, rel_tolerance_f16, tolerance_num, abs_tolerance_f16);
 }
 FIXTURE_DATA_TEST_CASE(RunLarge, NEGEMMFixture<half>, framework::DatasetMode::NIGHTLY, combine(datasets::LargeGEMMDataset(), framework::dataset::make("DataType",
                                                                                                DataType::F16)))
 {
     // Validate output
-    validate(Accessor(_target), _reference, tolerance_f);
+    validate(Accessor(_target), _reference, rel_tolerance_f16, tolerance_num, abs_tolerance_f16);
 }
 TEST_SUITE_END()
 #endif /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
diff --git a/tests/validation/NEON/Magnitude.cpp b/tests/validation/NEON/Magnitude.cpp
index 3b7562b61c..e1549a54b2 100644
--- a/tests/validation/NEON/Magnitude.cpp
+++ b/tests/validation/NEON/Magnitude.cpp
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -44,15 +44,6 @@ AbsoluteTolerance<T> tolerance(MagnitudeType magnitude_type)
 {
     return AbsoluteTolerance<T>((MagnitudeType::L1NORM == magnitude_type) ? 0 : 1);
 }
-
-#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-template <>
-AbsoluteTolerance<half_float::half> tolerance(MagnitudeType magnitude_type)
-{
-    return AbsoluteTolerance<half_float::half>((MagnitudeType::L1NORM == magnitude_type) ? half(0.0) : half(1.0));
-}
-#endif /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
-
 } // namespace
 
 TEST_SUITE(NEON)
@@ -88,36 +79,22 @@ template <typename T>
 using NEMagnitudeFixture = MagnitudeValidationFixture<Tensor, Accessor, NEMagnitude, T>;
 
 TEST_SUITE(S16)
-FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture<int16_t>, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S16)),
-                                                                                                                 framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                         framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture<int16_t>, framework::DatasetMode::PRECOMMIT, combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format", Format::S16)),
+                                                                                                         framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
+
 {
     // Validate output
     validate(Accessor(_target), _reference, tolerance<int16_t>(_magnitude_type));
 }
 
-FIXTURE_DATA_TEST_CASE(RunLarge, NEMagnitudeFixture<int16_t>, framework::DatasetMode::NIGHTLY, combine(combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S16)),
-                                                                                                               framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                       framework::dataset::make("UseFP16", false)))
+FIXTURE_DATA_TEST_CASE(RunLarge, NEMagnitudeFixture<int16_t>, framework::DatasetMode::NIGHTLY, combine(combine(datasets::Large2DShapes(), framework::dataset::make("Format", Format::S16)),
+                                                                                                       framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })))
 {
     // Validate output
     validate(Accessor(_target), _reference, tolerance<int16_t>(_magnitude_type));
 }
 TEST_SUITE_END() // S16
 
-#ifdef __ARM_FEATURE_FP16_VECTOR_ARITHMETIC
-TEST_SUITE(F16)
-FIXTURE_DATA_TEST_CASE(RunSmall, NEMagnitudeFixture<half_float::half>, framework::DatasetMode::PRECOMMIT, combine(combine(combine(datasets::Small2DShapes(), framework::dataset::make("Format",
-                                                                                                                  Format::S16)),
-                                                                                                                  framework::dataset::make("MagnitudeType", { MagnitudeType::L1NORM, MagnitudeType::L2NORM })),
-                                                                                                                  framework::dataset::make("UseFP16", true)))
-{
-    // Validate output
-    validate(Accessor(_target), _reference, tolerance<half_float::half>(_magnitude_type));
-}
-TEST_SUITE_END() // F16
-#endif           /* __ARM_FEATURE_FP16_VECTOR_ARITHMETIC */
-
 TEST_SUITE_END()
 TEST_SUITE_END()
 } // namespace validation
diff --git a/tests/validation/fixtures/MagnitudeFixture.h b/tests/validation/fixtures/MagnitudeFixture.h
index 1c529070a8..0930fb4117 100644
--- a/tests/validation/fixtures/MagnitudeFixture.h
+++ b/tests/validation/fixtures/MagnitudeFixture.h
@@ -1,5 +1,5 @@
 /*
- * Copyright (c) 2017 ARM Limited.
+ * Copyright (c) 2017-2018 ARM Limited.
  *
  * SPDX-License-Identifier: MIT
  *
@@ -43,9 +43,9 @@ class MagnitudeValidationFixture : public framework::Fixture
 {
 public:
     template <typename...>
-    void setup(TensorShape shape, Format format, MagnitudeType magnitude_type, bool use_fp16)
+    void setup(TensorShape shape, Format format, MagnitudeType magnitude_type)
     {
-        _target         = compute_target(shape, format, magnitude_type, use_fp16);
+        _target         = compute_target(shape, format, magnitude_type);
         _reference      = compute_reference(shape, format, magnitude_type);
         _magnitude_type = magnitude_type;
     }
@@ -57,7 +57,7 @@ protected:
         library->fill_tensor_uniform(tensor, seed_offset);
     }
 
-    TensorType compute_target(const TensorShape &shape, Format format, MagnitudeType magnitude_type, bool use_fp16)
+    TensorType compute_target(const TensorShape &shape, Format format, MagnitudeType magnitude_type)
     {
         DataType data_type = data_type_from_format(format);
 
@@ -73,7 +73,7 @@ protected:
 
         // Create and configure function
         FunctionType magnitude;
-        magnitude.configure(&src1, &src2, &dst, magnitude_type, use_fp16);
+        magnitude.configure(&src1, &src2, &dst, magnitude_type);
 
         ARM_COMPUTE_EXPECT(src1.info()->is_resizable(), framework::LogLevel::ERRORS);
         ARM_COMPUTE_EXPECT(src2.info()->is_resizable(), framework::LogLevel::ERRORS);