[Piglit] [PATCH 2/3] cl: Add support for OpenCV unit tests

[Dylan Baker]

I'm going to apologize in advance, I got a little excited about actually being 
able to review something, and might have gone a little overboard.

On Wednesday, January 15, 2014 12:13:06 PM Tom Stellard wrote:
> From: Tom Stellard <thomas.stellard at amd.com>
> 
> This enables piglit to run and interpret the results from OpenCV's
> gtest based opencv_test_ocl program.
> 
> This patch adds two new CMake configuration variables:
> 
> OPENCL_OpenCVTestOCL_BINDIR: You can use this variable to enable
> the OpenCV tests by setting it to the full path of the
> bin directory in your opencv build tree (e.g. /home/user/opencv/build/bin).
> 
> OPENCL_OpenCVTestOCL_WORKDIR: (Optional)If you don't want to use
> OpenCV's default work directory, you can use this variable to specify
> a different one.
> ---
>  CMakeLists.txt                                |  9 ++-
>  tests/all_cl.py                               |  4 ++
>  tests/cl/program/execute/opencv-merge-hist.cl | 98
> --------------------------- tests/ocl_config.py.in                        |
> 28 ++++++++
>  tests/opencv.py                               | 77 +++++++++++++++++++++
>  5 files changed, 117 insertions(+), 99 deletions(-)
>  delete mode 100644 tests/cl/program/execute/opencv-merge-hist.cl
>  create mode 100644 tests/ocl_config.py.in
>  create mode 100644 tests/opencv.py
> 
> diff --git a/CMakeLists.txt b/CMakeLists.txt
> index 1befffb..37eac65 100644
> --- a/CMakeLists.txt
> +++ b/CMakeLists.txt
> @@ -92,7 +92,10 @@ if(PIGLIT_BUILD_GLES3_TESTS AND NOT PIGLIT_USE_WAFFLE)
>  endif(PIGLIT_BUILD_GLES3_TESTS AND NOT PIGLIT_USE_WAFFLE)
> 
>  if(PIGLIT_BUILD_CL_TESTS)
> -	find_package(OpenCL REQUIRED)
> +        find_package(OpenCL REQUIRED)
> +        set(OPENCL_OpenCVTestOCL_BINDIR "" CACHE STRING "Directory with the
> opencv_ocl_test program" ) +        set(OPENCL_OpenCVTestOCL_WORKDIR ""
> CACHE STRING
> +            "Directory with the opencv test resources (This is not needed
> if your OpenCV build directory was $(opencv_src_dir)/build )" )
> endif(PIGLIT_BUILD_CL_TESTS)
> 
>  IF(${CMAKE_SYSTEM_NAME} MATCHES "Linux")
> @@ -356,6 +359,10 @@ configure_file(
>  	"${piglit_SOURCE_DIR}/tests/util/config.h.in"
>  	"${piglit_BINARY_DIR}/tests/util/config.h"
>  )
> +configure_file(
> +	"${piglit_SOURCE_DIR}/tests/ocl_config.py.in"
> +	"${piglit_BINARY_DIR}/tests/ocl_config.py"
> +)
> 
>  include(cmake/piglit_util.cmake)
>  include(cmake/piglit_glapi.cmake)
> diff --git a/tests/all_cl.py b/tests/all_cl.py
> index a7d7106..b9c112a 100644
> --- a/tests/all_cl.py
> +++ b/tests/all_cl.py
> @@ -7,6 +7,8 @@ __all__ = ['profile']
>  import os
>  import os.path as path
> 
> +from tests.opencv import add_opencv_tests
> +
>  from framework.core import Group, TestProfile
>  from framework.exectest import PlainExecTest
> 
> @@ -122,3 +124,5 @@ add_program_test_dir(program_execute_builtin,
> 'generated_tests/cl/builtin/relati program_execute_store = Group()
>  program["Execute"]["Store"] = program_execute_store
>  add_program_test_dir(program_execute_store, 'generated_tests/cl/store')
> +
> +add_opencv_tests(profile)
> diff --git a/tests/cl/program/execute/opencv-merge-hist.cl
> b/tests/cl/program/execute/opencv-merge-hist.cl deleted file mode 100644
> index 8dccf24..0000000
> --- a/tests/cl/program/execute/opencv-merge-hist.cl
> +++ /dev/null
> @@ -1,98 +0,0 @@
> -/*!
> -[config]
> -name: OpenCV merge-hist
> -kernel_name: merge_hist
> -dimensions: 1
> -global_size: 65536 1 1
> -local_size: 256 1 1
> -
> -# This test checks that the loop:
> -# for(int stride = HISTOGRAM256_WORK_GROUP_SIZE /2; stride > 0; stride >>=
> 1) -# in this kernel is unrolled correctly or not unrolled at all.  The
> Clang -# OpenCL frontend was unrolling this in way that created illegal
> uses of the -# barrier() builtin which resulted in GPU hangs on r600g.
> -[test]
> -arg_in:  0 buffer int[65536] repeat 0
> -arg_out: 1 buffer int[256]   repeat 0
> -arg_in:  2 int 256
> -
> -!*/
> -
> -// The kernel in this test is taken from the opencv library (opencv.org)
> -// File: modules/ocl/src/opencl/imgproc_histogram.cl
> -//
> -//                           License Agreement
> -//                For Open Source Computer Vision Library
> -//
> -// Copyright (C) 2010-2012, Institute Of Software Chinese Academy Of
> Science, all rights reserved. -// Copyright (C) 2010-2012, Advanced Micro
> Devices, Inc., all rights reserved. -// Copyright (C) 2010-2012,
> Multicoreware, Inc., all rights reserved. -// Third party copyrights are
> property of their respective owners. -//
> -// @Authors
> -//    Niko Li, newlife20080214 at gmail.com
> -//    Jia Haipeng, jiahaipeng95 at gmail.com
> -//    Xu Pang, pangxu010 at 163.com
> -//    Wenju He, wenju at multicorewareinc.com
> -// Redistribution and use in source and binary forms, with or without
> modification, -// are permitted provided that the following conditions are
> met:
> -//
> -//   * Redistribution's of source code must retain the above copyright
> notice, -//     this list of conditions and the following disclaimer.
> -//
> -//   * Redistribution's in binary form must reproduce the above copyright
> notice, -//     this list of conditions and the following disclaimer in the
> documentation -//     and/or other GpuMaterials provided with the
> distribution.
> -//
> -//   * The name of the copyright holders may not be used to endorse or
> promote products -//     derived from this software without specific prior
> written permission. -//
> -// This software is provided by the copyright holders and contributors as
> is and -// any express or implied warranties, including, but not limited
> to, the implied -// warranties of merchantability and fitness for a
> particular purpose are disclaimed. -// In no event shall the Intel
> Corporation or contributors be liable for any direct, -// indirect,
> incidental, special, exemplary, or consequential damages -// (including,
> but not limited to, procurement of substitute goods or services; -// loss
> of use, data, or profits; or business interruption) however caused -// and
> on any theory of liability, whether in contract, strict liability, -// or
> tort (including negligence or otherwise) arising in any way out of -// the
> use of this software, even if advised of the possibility of such damage.
> -//
> -//
> -
> -
> -#define PARTIAL_HISTOGRAM256_COUNT     (256)
> -#define HISTOGRAM256_BIN_COUNT         (256)
> -
> -#define HISTOGRAM256_WORK_GROUP_SIZE     (256)
> -#define HISTOGRAM256_LOCAL_MEM_SIZE      (HISTOGRAM256_BIN_COUNT)
> -
> -#define NBANKS (16)
> -#define NBANKS_BIT (4)
> -
> -
> -__kernel __attribute__((reqd_work_group_size(256,1,1)))void
> merge_hist(__global int* buf, -                __global int* hist,
> -                int src_step)
> -{
> -    int lx = get_local_id(0);
> -    int gx = get_group_id(0);
> -
> -    int sum = 0;
> -
> -    for(int i = lx; i < PARTIAL_HISTOGRAM256_COUNT; i +=
> HISTOGRAM256_WORK_GROUP_SIZE) -        sum += buf[ mad24(i, src_step, gx)];
> -
> -    __local int data[HISTOGRAM256_WORK_GROUP_SIZE];
> -    data[lx] = sum;
> -
> -    for(int stride = HISTOGRAM256_WORK_GROUP_SIZE /2; stride > 0; stride
> >>= 1) -    {
> -        barrier(CLK_LOCAL_MEM_FENCE);
> -        if(lx < stride)
> -            data[lx] += data[lx + stride];
> -    }
> -
> -    if(lx == 0)
> -        hist[gx] = data[0];
> -}
> diff --git a/tests/ocl_config.py.in b/tests/ocl_config.py.in
> new file mode 100644
> index 0000000..8cb90a8
> --- /dev/null
> +++ b/tests/ocl_config.py.in
> @@ -0,0 +1,28 @@
> +#!/usr/bin/env python
> +#
> +# Copyright 2014 Advanced Micro Devices, Inc.
> +#
> +# 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 (including the next
> +# paragraph) 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.
> +#
> +# Authors: Tom Stellard <thomas.stellard at amd.com>
> +#
> +
> +opencv_test_ocl_bin = '${OPENCL_OpenCVTestOCL_BINDIR}'
> +opencv_test_ocl_workdir = '${OPENCL_OpenCVTestOCL_WORKDIR}'

These are toplevel functions, they should be all capitalized.
ie: OPENCV_TEST_OCL_BIN

I think we would be better served using python's ConfigParser module and 
having a piglit configuration file. That might be a change for a separate 
series.

> diff --git a/tests/opencv.py b/tests/opencv.py
> new file mode 100644
> index 0000000..7dd06bb
> --- /dev/null
> +++ b/tests/opencv.py
> @@ -0,0 +1,77 @@
> +#!/usr/bin/env python
> +#
> +# Copyright 2014 Advanced Micro Devices, Inc.
> +#
> +# 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 (including the next
> +# paragraph) 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.
> +#
> +# Authors: Tom Stellard <thomas.stellard at amd.com>
> +#
> +
> +import os
> +import sys
> +import re
> +import string
> +
> +from os import path

This is a minor nit,  but could you put from "os import path" above the blank 
line?

> +from framework.exectest import ExecTest
> +from framework.core import TestProfile, testBinDir, Group
> +from framework.gtest import GTest
> +from tests.ocl_config import opencv_test_ocl_bin, opencv_test_ocl_workdir
> +
> +opencv_test_ocl = path.join(opencv_test_ocl_bin, 'opencv_test_ocl')
> +
> +class OpenCVTest(GTest):
> +    def __init__(self, testname):
> +        options = [opencv_test_ocl, '--gtest_filter=' 
> +        testname, '--gtest_color=no'] 
> +        if opencv_test_ocl_workdir != '':
> +            options.append('-w ' + opencv_test_ocl_workdir)

concatenation is expensive in python, using str.format is the generally 
accepted way to handle this.
options.append('-w {}'.format(opencv_test_ocl_workdir))

> +        ExecTest.__init__(self, options)
> +
> +
> +def add_opencv_tests(profile, individual = False):
> +    if opencv_test_ocl_bin == '':
> +        return
> +
> +    opencv = Group()
> +    tests = os.popen(opencv_test_ocl + " --gtest_list_tests").read()

os.popen is also deprecated, and again removed in python3
use the subprocess module instead. Note that it uses a list for getting 
arguments rather than a formatted string:
tests = subprocess.check_output([opencv_test_ocl, '--gtest_list_tests'])

> +    test_list = string.split(tests, '\n')
string is deprecated, and the str type provides what you want anyway:
test_list = tests.splitlines()

> +    group_name = ''
> +    for line in test_list:
> +        #Test groups names start at the beginning of the line and end with
> '.' +        m = re.match('([^.]+\.)$', line)
> +        if m is not None:

the python idiom is "if m:" or better yet "if re.match(..., line):"

> +            group_name = m.group(1)
> +            group_desc = group_name[:-1]
> +            if individual:
> +                opencv_group = Group()

This is just a suggestion but,
We have been making a push to move away from having groups, and instead just 
defining tests in the profile.test_lists dictionary:
profile['opencv/subgroup/test'] = OpenCVTest()
This is because working with nested groups is hard, and we flatten them into 
that dictionary form before working with them anyway.

> +                opencv[group_desc] = opencv_group
> +            else:
> +                opencv[group_desc] = OpenCVTest(group_name + "*")

see my other comment about concatenation

> +            continue
> +
> +        if not individual:
> +            continue
> +
> +        # Test names are indent by 2 spaces
> +        m = re.match('  (.+)', line)
> +        if m is not None:

see my previous comment about if.

> +            test_name = m.group(1)
> +            opencv_group[test_name] = OpenCVTest(group_name + test_name)
> +    profile.tests['opencv'] = opencv
-------------- next part --------------
A non-text attachment was scrubbed...
Name: signature.asc
Type: application/pgp-signature
Size: 490 bytes
Desc: This is a digitally signed message part.
URL: <http://lists.freedesktop.org/archives/piglit/attachments/20140115/4cd83280/attachment.pgp>