[Mesa-dev] [PATCH] cl: Add support for OpenCV unit tests v2
Dylan Baker
baker.dylan.c at gmail.com
Fri Jan 17 15:04:46 PST 2014
I think you meant to ping the piglit list not mesa dev :)
In general this looks a lot better, I've got a few little things for you
On Friday, January 17, 2014 02:30:13 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.
>
> v2:
> - Python code cleanups
>
> XXX: opencv
> ---
> CMakeLists.txt | 9 ++-
> tests/all_cl.py | 4 ++
> tests/cl/program/execute/opencv-merge-hist.cl | 98
> --------------------------- tests/ocl_config.py.in |
> 31 +++++++++
> tests/opencv.py | 76 +++++++++++++++++++++
> 5 files changed, 119 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..7ba515c
> --- /dev/null
> +++ b/tests/ocl_config.py.in
> @@ -0,0 +1,31 @@
> +#!/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>
> +#
> +
> +# TODO: Move this into a global piglit config file and use python's
> +# ConfigParser
> +
> +OPENCV_TEST_OCL_BINDIR = '${OPENCL_OpenCVTestOCL_BINDIR}'
> +OPENCV_TEST_OCL_WORKDIR = '${OPENCL_OpenCVTestOCL_WORKDIR}'
> diff --git a/tests/opencv.py b/tests/opencv.py
> new file mode 100644
> index 0000000..746a25b
> --- /dev/null
> +++ b/tests/opencv.py
> @@ -0,0 +1,76 @@
> +#!/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
> +import subprocess
> +from os import path
> +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
the imports on the last line are wrong, they should be all caps, and
OPENCV_TEST_OCL_BIN does not match what you have in ocl_config.py.in. There
are a few other cases of this
Also, you can remove the following imports you're not using: os, sys, string,
TestProfile, testBinDir, Exectest
> +
> +opencv_test_ocl = path.join(opencv_test_ocl_bin, 'opencv_test_ocl')
here's one of those cases
> +
> +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 {}'.format(opencv_test_ocl_workdir))
here's another of those cases
> + ExecTest.__init__(self, options)
> +
> +
> +def add_opencv_tests(profile, individual = False):
> + if opencv_test_ocl_bin == '':
and here
> + return
> +
> + opencv = Group()
> + tests = subprocess.check_output([opencv_test_ocl,
> '--gtest_list_tests']) + test_list = tests.splitlines()
> + group_name = ''
> + full_test_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:
> + group_name = m.group(1)
> + group_desc = group_name[:-1]
> + full_test_name = 'opencv/{}'.format(group_desc)
> + if not individual:
> + profile.tests[full_test_name] =
> OpenCVTest('{}*'.format(group_name)) + continue
> +
> + if not individual:
> + continue
> +
> + # Test names are indent by 2 spaces
> + m = re.match(' (.+)', line)
> + if m:
> + test_name = m.group(1)
> + profile.tests['{}/{}'.format(full_test_name,test_name)] = \
> + OpenCVTest('{}{}'.format(group_name
you only need one level of indent here beneath profile
profile.tests['{}/{}'.format(full_test_name,test_name)] = \
OpenCVTest('{}{}'.format(group_name
> ,test_name))
The only other thing is that this file should probably be in framework, not
tests. the only python that should be in tests are profiles, this is more test
classes.
with those changes you have my r-b:
Reviewed-by: Dylan Baker <baker.dylan.c at gmail.com>
-------------- 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/mesa-dev/attachments/20140117/4ee8d739/attachment-0001.pgp>
More information about the mesa-dev
mailing list