+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#ifndef OCL_HPP\r
-#define OCL_HPP\r
-\r
-#include "precomp.hpp"\r
-\r
-\r
-//Root level namespace\r
-namespace cv{\r
-\r
- //This namespace will contain all the source code for the OpenCL module\r
- namespace ocl{\r
-\r
- extern cl_context ocl_context;\r
- extern cl_command_queue ocl_cmd_queue;\r
- \r
- class OCL_EXPORTS OclMat{\r
- \r
- public:\r
- //! default constructor\r
- OclMat();\r
- \r
- //! constructs OclMat of the specified size and type (supported types are CV_8UC1, CV_8UC3, CV_16UC1, CV_16UC3, CV_64FC3 etc)\r
- OclMat(int rows, int cols, int type);\r
- OclMat(Size size, int type);\r
- //! constucts OclMat and fills it with the specified value _s.\r
- OclMat(int rows, int cols, int type, const Scalar& s);\r
- OclMat(Size size, int type, const Scalar& s);\r
- //! copy constructor\r
- OclMat(const OclMat& m);\r
-\r
- //! builds OclMat from Mat. Perfom blocking upload to device.\r
- explicit OclMat (const Mat& m);\r
-\r
- //! destructor - calls release()\r
- ~OclMat();\r
-\r
- //Releases the OpenCL context, command queue and the data buffer\r
- void release();\r
-\r
- //! assignment operators\r
- OclMat& operator = (const OclMat& m);\r
- //! assignment operator. Perfom blocking upload to device.\r
- OclMat& operator = (const Mat& m);\r
-\r
- //! sets every OclMatelement to s\r
- OclMat& operator = (const Scalar& s);\r
-\r
- //! sets some of the OclMat elements to s, according to the mask\r
- OclMat& setTo(const Scalar& s);\r
-\r
- //! pefroms blocking upload data to GpuMat.\r
- void upload(const cv::Mat& m);\r
-\r
- //! downloads data from device to host memory. Blocking calls.\r
- operator Mat();\r
- void download(cv::Mat& m);\r
-\r
- //! returns the size of element in bytes.\r
- size_t elemSize() const;\r
- //! returns the size of element channel in bytes.\r
- size_t elemSize1() const;\r
- //! returns element type, similar to CV_MAT_TYPE(cvMat->type)\r
- int type() const;\r
- //! returns element type, similar to CV_MAT_DEPTH(cvMat->type)\r
- int depth() const;\r
- //! returns element type, similar to CV_MAT_CN(cvMat->type)\r
- int channels() const;\r
- //! returns step/elemSize1()\r
- size_t step1() const;\r
- //! returns GpuMatrix size:\r
- // width == number of columns, height == number of rows\r
- Size size() const;\r
- //! returns true if OclMat data is NULL\r
- bool empty() const;\r
-\r
- int flags;\r
-\r
- //! the number of rows and columns\r
- int rows, cols;\r
- //! a distance between successive rows in bytes; includes the gap if any\r
- size_t step;\r
- //! pointer to the data of type cl_mem\r
- cl_mem data;\r
-\r
- int* refcount;\r
-\r
- //! allocates new OclMat data unless the OclMat already has specified size and type.\r
- // previous data is unreferenced if needed.\r
- void create(int rows, int cols, int type);\r
- void create(Size size, int type);\r
-\r
- void _upload(size_t size, void* src);\r
- void _download(size_t size, void* dst);\r
-\r
- };\r
-\r
- //Creates the OpenCL context and command queue\r
- OCL_EXPORTS void init();\r
-\r
- }\r
-}\r
-\r
-#endif
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other GpuMaterials provided with the distribution.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#ifndef OCL_UTIL_H\r
-#define OCL_UTIL_H\r
-\r
-#include "ocl.hpp"\r
-\r
-namespace cv{\r
- namespace ocl{\r
- namespace util{\r
-\r
- //Step 1: Get the target platform\r
- cl_platform_id GetOCLPlatform();\r
- \r
- //Step 2: Create a context\r
- int createContext(cl_context* context, cl_command_queue* cmd_queue, bool hasGPU);\r
- }\r
- }\r
-}\r
-\r
-#endif
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other GpuMaterials provided with the distribution.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-/*\r
-This header file will contain all the type, struct and enum definitions exclusively used by the OpenCL module\r
-*/\r
-\r
-#ifndef OCLDEFS_H\r
-#define OCLDEFS_H\r
-\r
-enum OclStatus{\r
-\r
- OCL_SUCCESS=0,\r
- OCL_ERROR=-100,\r
- OCL_SIZE_ERROR = -101\r
-\r
-};\r
-\r
-\r
-#endif
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-\r
-#include "ocl.hpp"\r
-#include <iostream>\r
-#include <fstream>\r
-using namespace std;\r
-\r
-namespace cv{\r
- namespace ocl{\r
-\r
- void writeBinaries(cl_program cpProgram)\r
- {\r
- ofstream myfile("kernel.ptx");\r
-\r
- cl_uint program_num_devices = 1;\r
- clGetProgramInfo(cpProgram, CL_PROGRAM_NUM_DEVICES, sizeof(cl_uint), &program_num_devices, NULL);\r
-\r
- if (program_num_devices == 0)\r
- return;\r
-\r
- size_t binaries_sizes[1];\r
-\r
- clGetProgramInfo(cpProgram, CL_PROGRAM_BINARY_SIZES, program_num_devices*sizeof(size_t), binaries_sizes, NULL);\r
-\r
- char **binaries = new char*[1];\r
- binaries[0] = 0;\r
-\r
- for (size_t i = 0; i < 1; i++)\r
- binaries[i] = new char[binaries_sizes[i]+1];\r
-\r
-\r
- clGetProgramInfo(cpProgram, CL_PROGRAM_BINARIES, program_num_devices*sizeof(size_t), binaries, NULL);\r
- \r
- if(myfile.is_open())\r
- {\r
- for (size_t i = 0; i < program_num_devices; i++)\r
- {\r
- myfile << binaries[i];\r
- }\r
- }\r
- myfile.close();\r
-\r
- for (size_t i = 0; i < program_num_devices; i++)\r
- delete [] binaries;\r
-\r
- delete [] binaries;\r
-}\r
-\r
- OCL_EXPORTS void add(const OclMat& a, const OclMat& b, OclMat& sum){\r
- \r
- if(a.rows != b.rows || a.cols != b.cols)\r
- return;\r
-\r
- int size = a.rows*a.cols;\r
-\r
- cl_program program;\r
- cl_kernel kernel;\r
- size_t global_work_size[1];\r
- size_t local_work_size[1];\r
- cl_uint size_ret = 0;\r
- cl_int err;\r
-\r
- const char* add_kernel_source[] = {\\r
- "__kernel void add_kernel (__global const uchar *a, __global const uchar* b, __global uchar* c)"\\r
- "{"\\r
- "int tid = get_global_id(0);"\\r
- "c[tid] = a[tid] + b[tid];"\\r
- "}"\r
- };\r
-\r
- program = clCreateProgramWithSource(ocl_context, 1, (const char**)&add_kernel_source, NULL, NULL);\r
-///////////////////\r
- writeBinaries(program);\r
-///////////////////\r
-\r
-\r
- //err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);\r
- /* FILE* fp = fopen("kernel.ptx", "r");\r
- fseek (fp , 0 , SEEK_END);\r
- const size_t lSize = ftell(fp);\r
- rewind(fp);\r
- unsigned char* buffer;\r
- buffer = (unsigned char*) malloc (lSize);\r
- fread(buffer, 1, lSize, fp);\r
- fclose(fp);\r
-\r
- size_t cb;\r
- err = clGetContextInfo(ocl_context, CL_CONTEXT_DEVICES, 0, NULL, &cb);\r
- cl_device_id *devices = (cl_device_id*)malloc(cb);\r
- clGetContextInfo(ocl_context, CL_CONTEXT_DEVICES, cb, devices, NULL);\r
-\r
- cl_int status;\r
- program = clCreateProgramWithBinary(ocl_context, 1, devices, \r
- &lSize, (const unsigned char**)&buffer, \r
- &status, &err);\r
- \r
- err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Build failed, check the program source...\n",err);\r
- return;\r
- }\r
-#endif\r
-\r
- //writeBinaries(program);\r
- \r
- kernel = clCreateKernel(program, "add_kernel", NULL);\r
-\r
- err = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &a.data);\r
- err = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *) &a.data);\r
- err = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *) &a.data);\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Failed at setting kernel arguments...\n",err);\r
- return;\r
- }\r
-#endif\r
-\r
- global_work_size[0] = size;\r
- local_work_size[0]= 1;\r
-\r
- err = clEnqueueNDRangeKernel(ocl_cmd_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Kernel execution failed...\n",err);\r
- return;\r
- }\r
-#endif\r
-\r
- clReleaseKernel(kernel);\r
- clReleaseProgram(program);*/\r
-\r
- }\r
- }\r
-}
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#include "ocl.hpp"\r
-\r
-cl_platform_id cv::ocl::util::GetOCLPlatform()\r
-{\r
- cl_platform_id pPlatforms[10] = { 0 };\r
- char pPlatformName[128] = { 0 };\r
-\r
- cl_uint uiPlatformsCount = 0;\r
- cl_int err = clGetPlatformIDs(10, pPlatforms, &uiPlatformsCount);\r
- for (cl_uint ui = 0; ui < uiPlatformsCount; ++ui)\r
- {\r
- err = clGetPlatformInfo(pPlatforms[ui], CL_PLATFORM_NAME, 128 * sizeof(char), pPlatformName, NULL);\r
- if ( err != CL_SUCCESS )\r
- {\r
- return NULL;\r
- }\r
-\r
- return pPlatforms[ui];\r
- }\r
-}\r
-\r
-int cv::ocl::util::createContext(cl_context* context, cl_command_queue* cmd_queue, bool hasGPU){\r
-\r
- size_t cb;\r
- cl_int err;\r
- cl_platform_id platform_id = cv::ocl::util::GetOCLPlatform();\r
-\r
- cl_context_properties context_properties[3] = {CL_CONTEXT_PLATFORM, (cl_context_properties)platform_id, NULL };\r
-\r
- //If the GPU is present, create the context on GPU\r
- if(hasGPU)\r
- *context = clCreateContextFromType(context_properties, CL_DEVICE_TYPE_GPU, NULL, NULL, NULL);\r
- else\r
- *context = clCreateContextFromType(context_properties, CL_DEVICE_TYPE_CPU, NULL, NULL, NULL);\r
-\r
- // get the list of devices associated with context\r
- err = clGetContextInfo(*context, CL_CONTEXT_DEVICES, 0, NULL, &cb);\r
-\r
- cl_device_id *devices = (cl_device_id*)malloc(cb);\r
-\r
- clGetContextInfo(*context, CL_CONTEXT_DEVICES, cb, devices, NULL);\r
-\r
- // create a command-queue\r
- *cmd_queue = clCreateCommandQueue(*context, devices[0], 0, NULL);\r
- free(devices);\r
-\r
- return CL_SUCCESS;\r
-}\r
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other GpuMaterials provided with the distribution.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#include "ocl.hpp"\r
-\r
-namespace cv{\r
-\r
- namespace ocl{\r
-\r
- cl_context ocl_context;\r
- cl_command_queue ocl_cmd_queue;\r
- bool initialized = false;\r
-\r
- inline OclMat::OclMat() : rows(0), cols(0), step(0), data(0), refcount(0) {}\r
-\r
- inline OclMat::OclMat(int _rows, int _cols, int _type): flags(0), rows(0), cols(0), step(0), data(0), refcount(0){\r
- \r
- if(_rows > 0 && _cols > 0)\r
- create(_rows, _cols, _type);\r
- }\r
-\r
- inline OclMat::OclMat(Size size, int _type): flags(0), rows(0), cols(0), step(0), data(0), refcount(0){\r
- \r
- if(size.height > 0 && size.width > 0)\r
- create(size, _type);\r
- }\r
-\r
- inline OclMat::OclMat(const OclMat& m) \r
- : flags(m.flags), rows(m.rows), cols(m.cols), step(m.step), data(m.data), refcount(m.refcount){\r
- \r
- if(refcount)\r
- CV_XADD(refcount, 1);\r
- }\r
-\r
- inline OclMat::OclMat(const Mat& m)\r
- : flags(0), rows(0), cols(0), step(0), data(0), refcount(0) { upload(m); }\r
-\r
- inline OclMat::~OclMat(){ release(); }\r
-\r
- void OclMat::_upload(size_t size, void* src){\r
-\r
- this->data = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, size, src, NULL);\r
- }\r
-\r
- void OclMat::_download(size_t size, void* dst){\r
- \r
- cl_int err = clEnqueueReadBuffer(ocl_cmd_queue, data, CL_TRUE, 0, size, dst, 0, NULL, NULL);\r
- }\r
-\r
-\r
- void OclMat::release(){\r
- \r
- if( refcount && CV_XADD(refcount, -1) == 1 )\r
- {\r
- free(refcount);\r
- clReleaseMemObject(data);\r
- }\r
- clReleaseMemObject(data);\r
- data = 0;\r
- step = rows = cols = 0;\r
- refcount = 0;\r
- }\r
-\r
- void OclMat::upload(const Mat& m){\r
- \r
- create(m.rows, m.cols, m.type());\r
- int ch = channels();\r
- int d = elemSize();\r
- size_t s = rows*cols*ch*d;\r
- if(initialized)\r
- this->_upload(s, m.data);\r
- else{\r
- init();\r
- this->_upload(s, m.data);\r
-#ifdef _DEBUG\r
- printf("Context and Command queues not initialized. First call cv::ocl::init()");\r
-#endif\r
- }\r
- }\r
-\r
- void OclMat::download(Mat& m){\r
- \r
- size_t s = rows*cols*channels()*elemSize();\r
- m.create(rows, cols, type());\r
- if(initialized){\r
- this->_download(s, m.data);\r
- }\r
-\r
- else{\r
- init();\r
- this->_download(s, m.data);\r
-#ifdef _DEBUG\r
- printf("Context and Command queues not initialized. First call cv::ocl::init()");\r
-#endif\r
-\r
- }\r
-\r
- }\r
-\r
- void init(){\r
-\r
- if(!initialized){\r
- cv::ocl::util::createContext(&ocl_context, &ocl_cmd_queue, false);\r
- initialized = true;\r
- }\r
- }\r
-\r
- void OclMat::create(int _rows, int _cols, int _type){\r
- \r
- if(!initialized)\r
- init();\r
-\r
- _type &= TYPE_MASK;\r
- \r
- if( rows == _rows && cols == _cols && type() == _type && data )\r
- return;\r
- if( data )\r
- release();\r
- \r
- if( _rows > 0 && _cols > 0 ){\r
- flags = Mat::MAGIC_VAL + _type;\r
- rows = _rows;\r
- cols = _cols;\r
-\r
- size_t esz = elemSize();\r
- int ch = channels();\r
-\r
- step = esz*cols*ch;\r
-\r
- size_t size = esz*rows*cols*ch;\r
- data = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE, size, NULL, NULL);\r
-\r
- if (esz * cols == step)\r
- flags |= Mat::CONTINUOUS_FLAG;\r
-\r
- /*\r
- refcount = (int*)fastMalloc(sizeof(*refcount));\r
- *refcount = 1;\r
- */\r
- }\r
- }\r
-\r
- void OclMat::create(Size size, int _type){\r
- \r
- return create(size.height, size.width, _type);\r
- }\r
-\r
- inline OclMat::operator Mat()\r
- {\r
- Mat m;\r
- download(m);\r
- return m;\r
- }\r
-\r
- inline OclMat& OclMat::operator = (const OclMat& m)\r
- {\r
- if( this != &m )\r
- {\r
- if( m.refcount )\r
- CV_XADD(m.refcount, 1);\r
- release();\r
- flags = m.flags;\r
- rows = m.rows; cols = m.cols;\r
- step = m.step; data = m.data;\r
- data = m.data;\r
- refcount = m.refcount;\r
- }\r
- return *this;\r
- }\r
-\r
- inline OclMat::OclMat(int _rows, int _cols, int _type, const Scalar& _s)\r
- : flags(0), rows(0), cols(0), step(0), data(0), refcount(0)\r
- {\r
- if(_rows > 0 && _cols > 0)\r
- {\r
- create(_rows, _cols, _type);\r
- *this = _s;\r
- }\r
- }\r
-\r
- inline OclMat& OclMat::operator = (const Mat& m) { upload(m); return *this; }\r
-\r
- OclMat& OclMat::operator = (const Scalar& s)\r
- {\r
- setTo(s);\r
- return *this;\r
- }\r
-\r
- OclMat& OclMat::setTo(const Scalar& s){\r
-\r
- //if (s[0] == 0.0 && s[1] == 0.0 && s[2] == 0.0 && s[3] == 0.0)\r
- //{\r
- \r
- size_t sz = rows*cols*channels()*elemSize();\r
- //void* ptr = (void*)malloc(sz);\r
- //memset(ptr, s[0], sz);\r
- //clEnqueueWriteBuffer(ocl_cmd_queue, data, CL_TRUE, 0, sz, ptr, NULL, NULL, NULL);\r
- //free(ptr); ptr = 0;\r
- return *this;\r
- //}\r
- }\r
-\r
- inline size_t OclMat::elemSize() const{ return CV_ELEM_SIZE(flags); }\r
- inline size_t OclMat::elemSize1() const{ return CV_ELEM_SIZE1(flags); }\r
- inline int OclMat::type() const { return CV_MAT_TYPE(flags); }\r
- inline int OclMat::depth() const{ return CV_MAT_DEPTH(flags); }\r
- inline int OclMat::channels() const{ return CV_MAT_CN(flags); }\r
- inline size_t OclMat::step1() const{ return step/elemSize1(); }\r
- inline Size OclMat::size() const{ return Size(cols, rows); }\r
- inline bool OclMat::empty() const{ return data == 0; }\r
- }\r
-}\r
-\r
-
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other GpuMaterials provided with the distribution.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#include "ocl.hpp"\r
-\r
-namespace cv{\r
- namespace ocl{\r
-\r
- extern cl_context ocl_context;\r
- extern cl_command_queue ocl_cmd_queue;\r
-\r
- OCL_EXPORTS void opticalFlowLK(const OclMat& imgA, const OclMat& imgB, CvSize winSize, OclMat& velX, OclMat& velY){\r
- \r
- if(imgA.rows != imgB.rows || imgA.cols != imgB.cols)\r
- return;\r
-\r
- int size = imgA.rows*imgA.cols;\r
-\r
- cl_program program;\r
- cl_kernel kernel;\r
- size_t global_work_size[1];\r
- size_t local_work_size[1];\r
- cl_uint size_ret = 0;\r
- cl_int err;\r
-\r
- \r
- //, __global const ushort* imWidth, __global const ushort* hradius\r
- const char* of_kernel_source[] = {\\r
- "__kernel void derivatives (__global const uchar *imgA, __global const uchar* imgB, __global float* fx, __global float* fy, __global float* ft, __global float* u, __global float* v)"\\r
- "{"\\r
- "int tid = get_global_id(0);"\\r
- "float Ix = 0.0f, Iy = 0.0f, It = 0.0f;"\\r
- "ushort imageWidth = 1024;"\\r
- "ushort half_radius = 1;"\\r
- "Ix = ((imgA[tid+1] - imgA[tid] + imgB[tid+1] - imgB[tid] )/2);"\\r
- "Iy = ((imgA[tid + imageWidth] - imgA[tid] + imgB[tid + imageWidth] - imgB[tid])/2);" \r
- "It = imgB[tid] - imgA[tid];"\\r
- "fx[tid] = Ix;"\\r
- "fy[tid] = Iy;"\\r
- "ft[tid] = It;"\\r
- "__local float s_data[3];"\\r
- "float A = 0.0f, B = 0.0f, C = 0.0f, D = 0.0f, E = 0.0f;"\\r
- "short i = 0;"\\r
- "for (i = -half_radius; i <= half_radius; i++){"\\r
- "for (short j = -half_radius; j <= half_radius; j++){"\\r
- "s_data[0] = fx[tid + i + j*imageWidth];"\\r
- "s_data[1] = fy[tid + i + j*imageWidth];"\\r
- "s_data[2] = ft[tid + i + j*imageWidth];"\\r
- "A = A + powf(s_data[0],2);"\\r
- "B = B + powf(s_data[1],2);"\\r
- "C = C + s_data[0] * s_data[1];"\\r
- "D = D + s_data[0] * s_data[2];"\\r
- "E = E + s_data[1] * s_data[2];"\\r
- "}}"\\r
- "u[tid] = (D*B - E*C)/(A*B - C*C);"\\r
- "v[tid] = (E*A - D*C)/(A*B - C*C);"\\r
- "}"\r
- };\r
-\r
- //program = clCreateProgramWithSource(imgA.ocl_context, 1, (const char**)&of_kernel_source, NULL, NULL);\r
- program = clCreateProgramWithSource(ocl_context, 1, (const char**)&of_kernel_source, NULL, NULL);\r
-\r
- err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);\r
-\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Build failed, check the program source...\n",err);\r
- return;\r
- }\r
-#endif\r
-\r
- kernel = clCreateKernel(program, "derivatives", NULL);\r
-\r
- //Creating additional temporary buffers fx, fy and ft\r
- //cl_mem fx = clCreateBuffer(imgA.ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
- //cl_mem fy = clCreateBuffer(imgA.ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
- //cl_mem ft = clCreateBuffer(imgA.ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
-\r
- cl_mem fx = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
- cl_mem fy = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
- cl_mem ft = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE, size*sizeof(cl_float), NULL, NULL);\r
-\r
- //Creating variables for imageWidth and half_window\r
- ushort x_radius = winSize.width/2;\r
- ushort cols = (ushort)imgA.cols;\r
- \r
- //cl_mem imageWidth = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(cl_ushort), &cols, NULL);\r
- //cl_mem half_radius = clCreateBuffer(ocl_context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(cl_ushort), &x_radius, NULL);\r
-\r
-\r
- err = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &imgA.data);\r
- err = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *) &imgB.data);\r
- err = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *) &fx);\r
- err = clSetKernelArg(kernel, 3, sizeof(cl_mem), (void *) &fy);\r
- err = clSetKernelArg(kernel, 4, sizeof(cl_mem), (void *) &ft);\r
- err = clSetKernelArg(kernel, 5, sizeof(cl_mem), (void *) &velX.data);\r
- err = clSetKernelArg(kernel, 6, sizeof(cl_mem), (void *) &velY.data);\r
- //err = clSetKernelArg(kernel, 7, sizeof(cl_mem), (void *) &imageWidth);\r
- //err = clSetKernelArg(kernel, 8, sizeof(cl_mem), (void *) &half_radius);\r
-\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Failed at setting kernel arguments...\n",err);\r
- return;\r
- }\r
-#endif\r
-\r
- global_work_size[0] = size;\r
- local_work_size[0]= 1;\r
-\r
- //err = clEnqueueNDRangeKernel(imgA.ocl_cmd_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);\r
- err = clEnqueueNDRangeKernel(ocl_cmd_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);\r
-#ifdef _DEBUG\r
- if(err != CL_SUCCESS){\r
- printf("(Error code: %d)Kernel execution failed...\n",err);\r
- return;\r
- }\r
-#endif\r
- clReleaseMemObject(fx);\r
- clReleaseMemObject(fy);\r
- clReleaseMemObject(ft);\r
- //clReleaseMemObject(imageWidth);\r
- //clReleaseMemObject(half_radius);\r
- clReleaseKernel(kernel);\r
- clReleaseProgram(program);\r
-\r
- }\r
- }\r
-}\r
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// Intel License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000, Intel Corporation, all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other materials provided with the distribution.\r
-//\r
-// * The name of Intel Corporation may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#include "precomp.hpp"\r
-\r
-/* End of file. */
\ No newline at end of file
+++ /dev/null
-/*M///////////////////////////////////////////////////////////////////////////////////////\r
-//\r
-// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING.\r
-//\r
-// By downloading, copying, installing or using the software you agree to this license.\r
-// If you do not agree to this license, do not download, install,\r
-// copy or use the software.\r
-//\r
-//\r
-// License Agreement\r
-// For Open Source Computer Vision Library\r
-//\r
-// Copyright (C) 2000-2008, Intel Corporation, all rights reserved.\r
-// Copyright (C) 2009, Willow Garage Inc., all rights reserved.\r
-// Third party copyrights are property of their respective owners.\r
-//\r
-// Redistribution and use in source and binary forms, with or without modification,\r
-// are permitted provided that the following conditions are met:\r
-//\r
-// * Redistribution's of source code must retain the above copyright notice,\r
-// this list of conditions and the following disclaimer.\r
-//\r
-// * Redistribution's in binary form must reproduce the above copyright notice,\r
-// this list of conditions and the following disclaimer in the documentation\r
-// and/or other materials provided with the distribution.\r
-//\r
-// * The name of the copyright holders may not be used to endorse or promote products\r
-// derived from this software without specific prior written permission.\r
-//\r
-// This software is provided by the copyright holders and contributors "as is" and\r
-// any express or implied warranties, including, but not limited to, the implied\r
-// warranties of merchantability and fitness for a particular purpose are disclaimed.\r
-// In no event shall the Intel Corporation or contributors be liable for any direct,\r
-// indirect, incidental, special, exemplary, or consequential damages\r
-// (including, but not limited to, procurement of substitute goods or services;\r
-// loss of use, data, or profits; or business interruption) however caused\r
-// and on any theory of liability, whether in contract, strict liability,\r
-// or tort (including negligence or otherwise) arising in any way out of\r
-// the use of this software, even if advised of the possibility of such damage.\r
-//\r
-//M*/\r
-\r
-#ifndef PRECOMP_H__\r
-#define PRECOMP_H__\r
-\r
-#define OCL_EXPORTS __declspec(dllexport)\r
-\r
-#include <exception>\r
-\r
-#include "opencv/highgui.h"\r
-#include "opencv/cv.h"\r
-#include <CL/cl.h> //General OpenCL include file. Current version is written for Intel architecture using Intel's OpenCL SDK\r
-#include "ocl.hpp"\r
-#include "oclmat.hpp"\r
-#include "ocldefs.h"\r
-#include "ocl_util.h"\r
-\r
-#ifdef _DEBUG\r
- #include <stdio.h>\r
-#endif\r
-\r
-#endif
\ No newline at end of file
projects / platform / upstream / opencv.git / commitdiff