From fd7bf0b76634396e14596ab4e82fb95eeb99811e Mon Sep 17 00:00:00 2001 From: Vladislav Vinogradov Date: Fri, 15 Mar 2013 14:09:39 +0400 Subject: [PATCH] moved SURF_GPU and VIBE to gpunonfree module --- modules/gpu/CMakeLists.txt | 2 +- .../gpu/doc/feature_detection_and_description.rst | 103 -------- modules/gpu/doc/video.rst | 71 ----- modules/gpu/include/opencv2/gpu/gpu.hpp | 111 -------- modules/gpu/perf/perf_video.cpp | 65 ----- modules/gpu/test/test_bgfg.cpp | 41 --- modules/gpunonfree/CMakeLists.txt | 31 +++ .../doc/feature_detection_and_description.rst | 107 ++++++++ modules/gpunonfree/doc/gpunonfree.rst | 9 + modules/gpunonfree/doc/video.rst | 79 ++++++ .../include/opencv2/gpunonfree/gpunonfree.hpp | 167 ++++++++++++ modules/gpunonfree/perf/perf_gpunonfree.cpp | 118 +++++++++ modules/gpunonfree/perf/perf_main.cpp | 5 + modules/gpunonfree/perf/perf_precomp.cpp | 1 + modules/gpunonfree/perf/perf_precomp.hpp | 31 +++ modules/{gpu => gpunonfree}/src/cuda/surf.cu | 30 +++ .../bgfg_vibe.cu => gpunonfree/src/cuda/vibe.cu} | 14 +- modules/gpunonfree/src/precomp.cpp | 45 ++++ modules/gpunonfree/src/precomp.hpp | 71 +++++ modules/{gpu => gpunonfree}/src/surf.cpp | 0 .../src/bgfg_vibe.cpp => gpunonfree/src/vibe.cpp} | 0 modules/gpunonfree/test/test_gpunonfree.cpp | 286 +++++++++++++++++++++ modules/gpunonfree/test/test_main.cpp | 119 +++++++++ modules/gpunonfree/test/test_precomp.cpp | 42 +++ modules/gpunonfree/test/test_precomp.hpp | 69 +++++ modules/stitching/CMakeLists.txt | 2 +- .../include/opencv2/stitching/detail/matchers.hpp | 6 +- modules/stitching/src/matchers.cpp | 2 +- modules/stitching/src/precomp.hpp | 3 + modules/stitching/src/stitcher.cpp | 2 +- samples/cpp/CMakeLists.txt | 11 +- samples/cpp/stitching_detailed.cpp | 2 +- samples/gpu/CMakeLists.txt | 9 +- samples/gpu/bgfg_segm.cpp | 37 ++- samples/gpu/driver_api_stereo_multi.cpp | 3 - samples/gpu/performance/CMakeLists.txt | 8 + samples/gpu/performance/tests.cpp | 10 +- samples/gpu/stereo_multi.cpp | 3 - samples/gpu/surf_keypoint_matcher.cpp | 15 ++ 39 files changed, 1317 insertions(+), 413 deletions(-) create mode 100644 modules/gpunonfree/CMakeLists.txt create mode 100644 modules/gpunonfree/doc/feature_detection_and_description.rst create mode 100644 modules/gpunonfree/doc/gpunonfree.rst create mode 100644 modules/gpunonfree/doc/video.rst create mode 100644 modules/gpunonfree/include/opencv2/gpunonfree/gpunonfree.hpp create mode 100644 modules/gpunonfree/perf/perf_gpunonfree.cpp create mode 100644 modules/gpunonfree/perf/perf_main.cpp create mode 100644 modules/gpunonfree/perf/perf_precomp.cpp create mode 100644 modules/gpunonfree/perf/perf_precomp.hpp rename modules/{gpu => gpunonfree}/src/cuda/surf.cu (96%) rename modules/{gpu/src/cuda/bgfg_vibe.cu => gpunonfree/src/cuda/vibe.cu} (95%) create mode 100644 modules/gpunonfree/src/precomp.cpp create mode 100644 modules/gpunonfree/src/precomp.hpp rename modules/{gpu => gpunonfree}/src/surf.cpp (100%) rename modules/{gpu/src/bgfg_vibe.cpp => gpunonfree/src/vibe.cpp} (100%) create mode 100644 modules/gpunonfree/test/test_gpunonfree.cpp create mode 100644 modules/gpunonfree/test/test_main.cpp create mode 100644 modules/gpunonfree/test/test_precomp.cpp create mode 100644 modules/gpunonfree/test/test_precomp.hpp diff --git a/modules/gpu/CMakeLists.txt b/modules/gpu/CMakeLists.txt index 26bf624..05b65f2 100644 --- a/modules/gpu/CMakeLists.txt +++ b/modules/gpu/CMakeLists.txt @@ -3,7 +3,7 @@ if(ANDROID OR IOS) endif() set(the_description "GPU-accelerated Computer Vision") -ocv_add_module(gpu opencv_imgproc opencv_calib3d opencv_objdetect opencv_video opencv_nonfree opencv_photo opencv_legacy) +ocv_add_module(gpu opencv_imgproc opencv_calib3d opencv_objdetect opencv_video opencv_photo opencv_legacy) ocv_module_include_directories("${CMAKE_CURRENT_SOURCE_DIR}/src/cuda") diff --git a/modules/gpu/doc/feature_detection_and_description.rst b/modules/gpu/doc/feature_detection_and_description.rst index ec75248..191b0ad 100644 --- a/modules/gpu/doc/feature_detection_and_description.rst +++ b/modules/gpu/doc/feature_detection_and_description.rst @@ -5,109 +5,6 @@ Feature Detection and Description -gpu::SURF_GPU -------------- -.. ocv:class:: gpu::SURF_GPU - -Class used for extracting Speeded Up Robust Features (SURF) from an image. :: - - class SURF_GPU - { - public: - enum KeypointLayout - { - X_ROW = 0, - Y_ROW, - LAPLACIAN_ROW, - OCTAVE_ROW, - SIZE_ROW, - ANGLE_ROW, - HESSIAN_ROW, - ROWS_COUNT - }; - - //! the default constructor - SURF_GPU(); - //! the full constructor taking all the necessary parameters - explicit SURF_GPU(double _hessianThreshold, int _nOctaves=4, - int _nOctaveLayers=2, bool _extended=false, float _keypointsRatio=0.01f); - - //! returns the descriptor size in float's (64 or 128) - int descriptorSize() const; - - //! upload host keypoints to device memory - void uploadKeypoints(const vector& keypoints, - GpuMat& keypointsGPU); - //! download keypoints from device to host memory - void downloadKeypoints(const GpuMat& keypointsGPU, - vector& keypoints); - - //! download descriptors from device to host memory - void downloadDescriptors(const GpuMat& descriptorsGPU, - vector& descriptors); - - void operator()(const GpuMat& img, const GpuMat& mask, - GpuMat& keypoints); - - void operator()(const GpuMat& img, const GpuMat& mask, - GpuMat& keypoints, GpuMat& descriptors, - bool useProvidedKeypoints = false, - bool calcOrientation = true); - - void operator()(const GpuMat& img, const GpuMat& mask, - std::vector& keypoints); - - void operator()(const GpuMat& img, const GpuMat& mask, - std::vector& keypoints, GpuMat& descriptors, - bool useProvidedKeypoints = false, - bool calcOrientation = true); - - void operator()(const GpuMat& img, const GpuMat& mask, - std::vector& keypoints, - std::vector& descriptors, - bool useProvidedKeypoints = false, - bool calcOrientation = true); - - void releaseMemory(); - - // SURF parameters - double hessianThreshold; - int nOctaves; - int nOctaveLayers; - bool extended; - bool upright; - - //! max keypoints = keypointsRatio * img.size().area() - float keypointsRatio; - - GpuMat sum, mask1, maskSum, intBuffer; - - GpuMat det, trace; - - GpuMat maxPosBuffer; - }; - - -The class ``SURF_GPU`` implements Speeded Up Robust Features descriptor. There is a fast multi-scale Hessian keypoint detector that can be used to find the keypoints (which is the default option). But the descriptors can also be computed for the user-specified keypoints. Only 8-bit grayscale images are supported. - -The class ``SURF_GPU`` can store results in the GPU and CPU memory. It provides functions to convert results between CPU and GPU version ( ``uploadKeypoints``, ``downloadKeypoints``, ``downloadDescriptors`` ). The format of CPU results is the same as ``SURF`` results. GPU results are stored in ``GpuMat``. The ``keypoints`` matrix is :math:`\texttt{nFeatures} \times 7` matrix with the ``CV_32FC1`` type. - -* ``keypoints.ptr(X_ROW)[i]`` contains x coordinate of the i-th feature. -* ``keypoints.ptr(Y_ROW)[i]`` contains y coordinate of the i-th feature. -* ``keypoints.ptr(LAPLACIAN_ROW)[i]`` contains the laplacian sign of the i-th feature. -* ``keypoints.ptr(OCTAVE_ROW)[i]`` contains the octave of the i-th feature. -* ``keypoints.ptr(SIZE_ROW)[i]`` contains the size of the i-th feature. -* ``keypoints.ptr(ANGLE_ROW)[i]`` contain orientation of the i-th feature. -* ``keypoints.ptr(HESSIAN_ROW)[i]`` contains the response of the i-th feature. - -The ``descriptors`` matrix is :math:`\texttt{nFeatures} \times \texttt{descriptorSize}` matrix with the ``CV_32FC1`` type. - -The class ``SURF_GPU`` uses some buffers and provides access to it. All buffers can be safely released between function calls. - -.. seealso:: :ocv:class:`SURF` - - - gpu::FAST_GPU ------------- .. ocv:class:: gpu::FAST_GPU diff --git a/modules/gpu/doc/video.rst b/modules/gpu/doc/video.rst index fc5b1fb..284bb17 100644 --- a/modules/gpu/doc/video.rst +++ b/modules/gpu/doc/video.rst @@ -579,76 +579,6 @@ Releases all inner buffer's memory. -gpu::VIBE_GPU -------------- -.. ocv:class:: gpu::VIBE_GPU - -Class used for background/foreground segmentation. :: - - class VIBE_GPU - { - public: - explicit VIBE_GPU(unsigned long rngSeed = 1234567); - - void initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()); - - void operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()); - - void release(); - - ... - }; - -The class discriminates between foreground and background pixels by building and maintaining a model of the background. Any pixel which does not fit this model is then deemed to be foreground. The class implements algorithm described in [VIBE2011]_. - - - -gpu::VIBE_GPU::VIBE_GPU ------------------------ -The constructor. - -.. ocv:function:: gpu::VIBE_GPU::VIBE_GPU(unsigned long rngSeed = 1234567) - - :param rngSeed: Value used to initiate a random sequence. - -Default constructor sets all parameters to default values. - - - -gpu::VIBE_GPU::initialize -------------------------- -Initialize background model and allocates all inner buffers. - -.. ocv:function:: void gpu::VIBE_GPU::initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()) - - :param firstFrame: First frame from video sequence. - - :param stream: Stream for the asynchronous version. - - - -gpu::VIBE_GPU::operator() -------------------------- -Updates the background model and returns the foreground mask - -.. ocv:function:: void gpu::VIBE_GPU::operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()) - - :param frame: Next video frame. - - :param fgmask: The output foreground mask as an 8-bit binary image. - - :param stream: Stream for the asynchronous version. - - - -gpu::VIBE_GPU::release ----------------------- -Releases all inner buffer's memory. - -.. ocv:function:: void gpu::VIBE_GPU::release() - - - gpu::GMG_GPU ------------ .. ocv:class:: gpu::GMG_GPU @@ -1209,5 +1139,4 @@ Parse next video frame. Implementation must call this method after new frame was .. [MOG2001] P. KadewTraKuPong and R. Bowden. *An improved adaptive background mixture model for real-time tracking with shadow detection*. Proc. 2nd European Workshop on Advanced Video-Based Surveillance Systems, 2001 .. [MOG2004] Z. Zivkovic. *Improved adaptive Gausian mixture model for background subtraction*. International Conference Pattern Recognition, UK, August, 2004 .. [ShadowDetect2003] Prati, Mikic, Trivedi and Cucchiarra. *Detecting Moving Shadows...*. IEEE PAMI, 2003 -.. [VIBE2011] O. Barnich and M. Van D Roogenbroeck. *ViBe: A universal background subtraction algorithm for video sequences*. IEEE Transactions on Image Processing, 20(6) :1709-1724, June 2011 .. [GMG2012] A. Godbehere, A. Matsukawa and K. Goldberg. *Visual Tracking of Human Visitors under Variable-Lighting Conditions for a Responsive Audio Art Installation*. American Control Conference, Montreal, June 2012 diff --git a/modules/gpu/include/opencv2/gpu/gpu.hpp b/modules/gpu/include/opencv2/gpu/gpu.hpp index 802954c..bb74d1a 100644 --- a/modules/gpu/include/opencv2/gpu/gpu.hpp +++ b/modules/gpu/include/opencv2/gpu/gpu.hpp @@ -1557,82 +1557,6 @@ private: friend class CascadeClassifier_GPU_LBP; }; -////////////////////////////////// SURF ////////////////////////////////////////// - -class CV_EXPORTS SURF_GPU -{ -public: - enum KeypointLayout - { - X_ROW = 0, - Y_ROW, - LAPLACIAN_ROW, - OCTAVE_ROW, - SIZE_ROW, - ANGLE_ROW, - HESSIAN_ROW, - ROWS_COUNT - }; - - //! the default constructor - SURF_GPU(); - //! the full constructor taking all the necessary parameters - explicit SURF_GPU(double _hessianThreshold, int _nOctaves=4, - int _nOctaveLayers=2, bool _extended=false, float _keypointsRatio=0.01f, bool _upright = false); - - //! returns the descriptor size in float's (64 or 128) - int descriptorSize() const; - - //! upload host keypoints to device memory - void uploadKeypoints(const vector& keypoints, GpuMat& keypointsGPU); - //! download keypoints from device to host memory - void downloadKeypoints(const GpuMat& keypointsGPU, vector& keypoints); - - //! download descriptors from device to host memory - void downloadDescriptors(const GpuMat& descriptorsGPU, vector& descriptors); - - //! finds the keypoints using fast hessian detector used in SURF - //! supports CV_8UC1 images - //! keypoints will have nFeature cols and 6 rows - //! keypoints.ptr(X_ROW)[i] will contain x coordinate of i'th feature - //! keypoints.ptr(Y_ROW)[i] will contain y coordinate of i'th feature - //! keypoints.ptr(LAPLACIAN_ROW)[i] will contain laplacian sign of i'th feature - //! keypoints.ptr(OCTAVE_ROW)[i] will contain octave of i'th feature - //! keypoints.ptr(SIZE_ROW)[i] will contain size of i'th feature - //! keypoints.ptr(ANGLE_ROW)[i] will contain orientation of i'th feature - //! keypoints.ptr(HESSIAN_ROW)[i] will contain response of i'th feature - void operator()(const GpuMat& img, const GpuMat& mask, GpuMat& keypoints); - //! finds the keypoints and computes their descriptors. - //! Optionally it can compute descriptors for the user-provided keypoints and recompute keypoints direction - void operator()(const GpuMat& img, const GpuMat& mask, GpuMat& keypoints, GpuMat& descriptors, - bool useProvidedKeypoints = false); - - void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints); - void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints, GpuMat& descriptors, - bool useProvidedKeypoints = false); - - void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints, std::vector& descriptors, - bool useProvidedKeypoints = false); - - void releaseMemory(); - - // SURF parameters - double hessianThreshold; - int nOctaves; - int nOctaveLayers; - bool extended; - bool upright; - - //! max keypoints = min(keypointsRatio * img.size().area(), 65535) - float keypointsRatio; - - GpuMat sum, mask1, maskSum, intBuffer; - - GpuMat det, trace; - - GpuMat maxPosBuffer; -}; - ////////////////////////////////// FAST ////////////////////////////////////////// class CV_EXPORTS FAST_GPU @@ -2307,41 +2231,6 @@ private: GpuMat bgmodelUsedModes_; //keep track of number of modes per pixel }; -/*! - * The class implements the following algorithm: - * "ViBe: A universal background subtraction algorithm for video sequences" - * O. Barnich and M. Van D Roogenbroeck - * IEEE Transactions on Image Processing, 20(6) :1709-1724, June 2011 - */ -class CV_EXPORTS VIBE_GPU -{ -public: - //! the default constructor - explicit VIBE_GPU(unsigned long rngSeed = 1234567); - - //! re-initiaization method - void initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()); - - //! the update operator - void operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()); - - //! releases all inner buffers - void release(); - - int nbSamples; // number of samples per pixel - int reqMatches; // #_min - int radius; // R - int subsamplingFactor; // amount of random subsampling - -private: - Size frameSize_; - - unsigned long rngSeed_; - GpuMat randStates_; - - GpuMat samples_; -}; - /** * Background Subtractor module. Takes a series of images and returns a sequence of mask (8UC1) * images of the same size, where 255 indicates Foreground and 0 represents Background. diff --git a/modules/gpu/perf/perf_video.cpp b/modules/gpu/perf/perf_video.cpp index 61c2687..e9b38ed 100644 --- a/modules/gpu/perf/perf_video.cpp +++ b/modules/gpu/perf/perf_video.cpp @@ -819,71 +819,6 @@ PERF_TEST_P(Video_Cn, Video_MOG2GetBackgroundImage, } ////////////////////////////////////////////////////// -// VIBE - -PERF_TEST_P(Video_Cn, Video_VIBE, - Combine(Values("gpu/video/768x576.avi", "gpu/video/1920x1080.avi"), - GPU_CHANNELS_1_3_4)) -{ - const string inputFile = perf::TestBase::getDataPath(GET_PARAM(0)); - const int cn = GET_PARAM(1); - - cv::VideoCapture cap(inputFile); - ASSERT_TRUE(cap.isOpened()); - - cv::Mat frame; - cap >> frame; - ASSERT_FALSE(frame.empty()); - - if (cn != 3) - { - cv::Mat temp; - if (cn == 1) - cv::cvtColor(frame, temp, cv::COLOR_BGR2GRAY); - else - cv::cvtColor(frame, temp, cv::COLOR_BGR2BGRA); - cv::swap(temp, frame); - } - - if (PERF_RUN_GPU()) - { - cv::gpu::GpuMat d_frame(frame); - cv::gpu::VIBE_GPU vibe; - cv::gpu::GpuMat foreground; - - vibe(d_frame, foreground); - - for (int i = 0; i < 10; ++i) - { - cap >> frame; - ASSERT_FALSE(frame.empty()); - - if (cn != 3) - { - cv::Mat temp; - if (cn == 1) - cv::cvtColor(frame, temp, cv::COLOR_BGR2GRAY); - else - cv::cvtColor(frame, temp, cv::COLOR_BGR2BGRA); - cv::swap(temp, frame); - } - - d_frame.upload(frame); - - startTimer(); next(); - vibe(d_frame, foreground); - stopTimer(); - } - - GPU_SANITY_CHECK(foreground); - } - else - { - FAIL_NO_CPU(); - } -} - -////////////////////////////////////////////////////// // GMG DEF_PARAM_TEST(Video_Cn_MaxFeatures, string, MatCn, int); diff --git a/modules/gpu/test/test_bgfg.cpp b/modules/gpu/test/test_bgfg.cpp index aea6eb7..253ded2 100644 --- a/modules/gpu/test/test_bgfg.cpp +++ b/modules/gpu/test/test_bgfg.cpp @@ -323,47 +323,6 @@ INSTANTIATE_TEST_CASE_P(GPU_Video, MOG2, testing::Combine( WHOLE_SUBMAT)); ////////////////////////////////////////////////////// -// VIBE - -PARAM_TEST_CASE(VIBE, cv::gpu::DeviceInfo, cv::Size, MatType, UseRoi) -{ -}; - -GPU_TEST_P(VIBE, Accuracy) -{ - const cv::gpu::DeviceInfo devInfo = GET_PARAM(0); - cv::gpu::setDevice(devInfo.deviceID()); - const cv::Size size = GET_PARAM(1); - const int type = GET_PARAM(2); - const bool useRoi = GET_PARAM(3); - - const cv::Mat fullfg(size, CV_8UC1, cv::Scalar::all(255)); - - cv::Mat frame = randomMat(size, type, 0.0, 100); - cv::gpu::GpuMat d_frame = loadMat(frame, useRoi); - - cv::gpu::VIBE_GPU vibe; - cv::gpu::GpuMat d_fgmask = createMat(size, CV_8UC1, useRoi); - vibe.initialize(d_frame); - - for (int i = 0; i < 20; ++i) - vibe(d_frame, d_fgmask); - - frame = randomMat(size, type, 160, 255); - d_frame = loadMat(frame, useRoi); - vibe(d_frame, d_fgmask); - - // now fgmask should be entirely foreground - ASSERT_MAT_NEAR(fullfg, d_fgmask, 0); -} - -INSTANTIATE_TEST_CASE_P(GPU_Video, VIBE, testing::Combine( - ALL_DEVICES, - DIFFERENT_SIZES, - testing::Values(MatType(CV_8UC1), MatType(CV_8UC3), MatType(CV_8UC4)), - WHOLE_SUBMAT)); - -////////////////////////////////////////////////////// // GMG PARAM_TEST_CASE(GMG, cv::gpu::DeviceInfo, cv::Size, MatDepth, Channels, UseRoi) diff --git a/modules/gpunonfree/CMakeLists.txt b/modules/gpunonfree/CMakeLists.txt new file mode 100644 index 0000000..a975a61 --- /dev/null +++ b/modules/gpunonfree/CMakeLists.txt @@ -0,0 +1,31 @@ +if(ANDROID OR IOS) + ocv_module_disable(gpunonfree) +endif() + +set(the_description "GPU-accelerated Computer Vision (non free)") +ocv_add_module(gpunonfree opencv_gpu opencv_nonfree) +ocv_module_include_directories() + +if(HAVE_CUDA) + ocv_source_group("Src\\Cuda" GLOB "src/cuda/*.cu") + ocv_include_directories("${OpenCV_SOURCE_DIR}/modules/gpu/include" ${CUDA_INCLUDE_DIRS}) + ocv_warnings_disable(CMAKE_CXX_FLAGS -Wundef) + + file(GLOB lib_cuda "src/cuda/*.cu") + ocv_cuda_compile(cuda_objs ${lib_cuda}) + + set(cuda_link_libs ${CUDA_LIBRARIES} ${CUDA_npp_LIBRARY}) +else() + set(lib_cuda "") + set(cuda_objs "") + set(cuda_link_libs "") +endif() + +ocv_glob_module_sources(SOURCES ${lib_cuda} ${cuda_objs}) + +ocv_create_module(${cuda_link_libs}) +ocv_add_precompiled_headers(${the_module}) + +ocv_add_accuracy_tests() +ocv_add_perf_tests() + diff --git a/modules/gpunonfree/doc/feature_detection_and_description.rst b/modules/gpunonfree/doc/feature_detection_and_description.rst new file mode 100644 index 0000000..9cec66f --- /dev/null +++ b/modules/gpunonfree/doc/feature_detection_and_description.rst @@ -0,0 +1,107 @@ +Feature Detection and Description +================================= + +.. highlight:: cpp + + + +gpu::SURF_GPU +------------- +.. ocv:class:: gpu::SURF_GPU + +Class used for extracting Speeded Up Robust Features (SURF) from an image. :: + + class SURF_GPU + { + public: + enum KeypointLayout + { + X_ROW = 0, + Y_ROW, + LAPLACIAN_ROW, + OCTAVE_ROW, + SIZE_ROW, + ANGLE_ROW, + HESSIAN_ROW, + ROWS_COUNT + }; + + //! the default constructor + SURF_GPU(); + //! the full constructor taking all the necessary parameters + explicit SURF_GPU(double _hessianThreshold, int _nOctaves=4, + int _nOctaveLayers=2, bool _extended=false, float _keypointsRatio=0.01f); + + //! returns the descriptor size in float's (64 or 128) + int descriptorSize() const; + + //! upload host keypoints to device memory + void uploadKeypoints(const vector& keypoints, + GpuMat& keypointsGPU); + //! download keypoints from device to host memory + void downloadKeypoints(const GpuMat& keypointsGPU, + vector& keypoints); + + //! download descriptors from device to host memory + void downloadDescriptors(const GpuMat& descriptorsGPU, + vector& descriptors); + + void operator()(const GpuMat& img, const GpuMat& mask, + GpuMat& keypoints); + + void operator()(const GpuMat& img, const GpuMat& mask, + GpuMat& keypoints, GpuMat& descriptors, + bool useProvidedKeypoints = false, + bool calcOrientation = true); + + void operator()(const GpuMat& img, const GpuMat& mask, + std::vector& keypoints); + + void operator()(const GpuMat& img, const GpuMat& mask, + std::vector& keypoints, GpuMat& descriptors, + bool useProvidedKeypoints = false, + bool calcOrientation = true); + + void operator()(const GpuMat& img, const GpuMat& mask, + std::vector& keypoints, + std::vector& descriptors, + bool useProvidedKeypoints = false, + bool calcOrientation = true); + + void releaseMemory(); + + // SURF parameters + double hessianThreshold; + int nOctaves; + int nOctaveLayers; + bool extended; + bool upright; + + //! max keypoints = keypointsRatio * img.size().area() + float keypointsRatio; + + GpuMat sum, mask1, maskSum, intBuffer; + + GpuMat det, trace; + + GpuMat maxPosBuffer; + }; + + +The class ``SURF_GPU`` implements Speeded Up Robust Features descriptor. There is a fast multi-scale Hessian keypoint detector that can be used to find the keypoints (which is the default option). But the descriptors can also be computed for the user-specified keypoints. Only 8-bit grayscale images are supported. + +The class ``SURF_GPU`` can store results in the GPU and CPU memory. It provides functions to convert results between CPU and GPU version ( ``uploadKeypoints``, ``downloadKeypoints``, ``downloadDescriptors`` ). The format of CPU results is the same as ``SURF`` results. GPU results are stored in ``GpuMat``. The ``keypoints`` matrix is :math:`\texttt{nFeatures} \times 7` matrix with the ``CV_32FC1`` type. + +* ``keypoints.ptr(X_ROW)[i]`` contains x coordinate of the i-th feature. +* ``keypoints.ptr(Y_ROW)[i]`` contains y coordinate of the i-th feature. +* ``keypoints.ptr(LAPLACIAN_ROW)[i]`` contains the laplacian sign of the i-th feature. +* ``keypoints.ptr(OCTAVE_ROW)[i]`` contains the octave of the i-th feature. +* ``keypoints.ptr(SIZE_ROW)[i]`` contains the size of the i-th feature. +* ``keypoints.ptr(ANGLE_ROW)[i]`` contain orientation of the i-th feature. +* ``keypoints.ptr(HESSIAN_ROW)[i]`` contains the response of the i-th feature. + +The ``descriptors`` matrix is :math:`\texttt{nFeatures} \times \texttt{descriptorSize}` matrix with the ``CV_32FC1`` type. + +The class ``SURF_GPU`` uses some buffers and provides access to it. All buffers can be safely released between function calls. + +.. seealso:: :ocv:class:`SURF` diff --git a/modules/gpunonfree/doc/gpunonfree.rst b/modules/gpunonfree/doc/gpunonfree.rst new file mode 100644 index 0000000..4c2b71f --- /dev/null +++ b/modules/gpunonfree/doc/gpunonfree.rst @@ -0,0 +1,9 @@ +****************************************************** +gpunonfree. GPU-accelerated Computer Vision (non free) +****************************************************** + +.. toctree:: + :maxdepth: 1 + + feature_detection_and_description + video diff --git a/modules/gpunonfree/doc/video.rst b/modules/gpunonfree/doc/video.rst new file mode 100644 index 0000000..4fdfacc --- /dev/null +++ b/modules/gpunonfree/doc/video.rst @@ -0,0 +1,79 @@ +Video Analysis +============== + +.. highlight:: cpp + + + +gpu::VIBE_GPU +------------- +.. ocv:class:: gpu::VIBE_GPU + +Class used for background/foreground segmentation. :: + + class VIBE_GPU + { + public: + explicit VIBE_GPU(unsigned long rngSeed = 1234567); + + void initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()); + + void operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()); + + void release(); + + ... + }; + +The class discriminates between foreground and background pixels by building and maintaining a model of the background. Any pixel which does not fit this model is then deemed to be foreground. The class implements algorithm described in [VIBE2011]_. + + + +gpu::VIBE_GPU::VIBE_GPU +----------------------- +The constructor. + +.. ocv:function:: gpu::VIBE_GPU::VIBE_GPU(unsigned long rngSeed = 1234567) + + :param rngSeed: Value used to initiate a random sequence. + +Default constructor sets all parameters to default values. + + + +gpu::VIBE_GPU::initialize +------------------------- +Initialize background model and allocates all inner buffers. + +.. ocv:function:: void gpu::VIBE_GPU::initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()) + + :param firstFrame: First frame from video sequence. + + :param stream: Stream for the asynchronous version. + + + +gpu::VIBE_GPU::operator() +------------------------- +Updates the background model and returns the foreground mask + +.. ocv:function:: void gpu::VIBE_GPU::operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()) + + :param frame: Next video frame. + + :param fgmask: The output foreground mask as an 8-bit binary image. + + :param stream: Stream for the asynchronous version. + + + +gpu::VIBE_GPU::release +---------------------- +Releases all inner buffer's memory. + +.. ocv:function:: void gpu::VIBE_GPU::release() + + + + +.. [VIBE2011] O. Barnich and M. Van D Roogenbroeck. *ViBe: A universal background subtraction algorithm for video sequences*. IEEE Transactions on Image Processing, 20(6) :1709-1724, June 2011 diff --git a/modules/gpunonfree/include/opencv2/gpunonfree/gpunonfree.hpp b/modules/gpunonfree/include/opencv2/gpunonfree/gpunonfree.hpp new file mode 100644 index 0000000..40d9cfd --- /dev/null +++ b/modules/gpunonfree/include/opencv2/gpunonfree/gpunonfree.hpp @@ -0,0 +1,167 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000-2008, Intel Corporation, all rights reserved. +// Copyright (C) 2009, Willow Garage Inc., all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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. +// +//M*/ + +#ifndef __OPENCV_GPUNONFREE_HPP__ +#define __OPENCV_GPUNONFREE_HPP__ + +#ifndef SKIP_INCLUDES +#include +#endif + +#include "opencv2/gpu/gpu.hpp" + +namespace cv { namespace gpu { + +class CV_EXPORTS SURF_GPU +{ +public: + enum KeypointLayout + { + X_ROW = 0, + Y_ROW, + LAPLACIAN_ROW, + OCTAVE_ROW, + SIZE_ROW, + ANGLE_ROW, + HESSIAN_ROW, + ROWS_COUNT + }; + + //! the default constructor + SURF_GPU(); + //! the full constructor taking all the necessary parameters + explicit SURF_GPU(double _hessianThreshold, int _nOctaves=4, + int _nOctaveLayers=2, bool _extended=false, float _keypointsRatio=0.01f, bool _upright = false); + + //! returns the descriptor size in float's (64 or 128) + int descriptorSize() const; + + //! upload host keypoints to device memory + void uploadKeypoints(const std::vector& keypoints, GpuMat& keypointsGPU); + //! download keypoints from device to host memory + void downloadKeypoints(const GpuMat& keypointsGPU, std::vector& keypoints); + + //! download descriptors from device to host memory + void downloadDescriptors(const GpuMat& descriptorsGPU, std::vector& descriptors); + + //! finds the keypoints using fast hessian detector used in SURF + //! supports CV_8UC1 images + //! keypoints will have nFeature cols and 6 rows + //! keypoints.ptr(X_ROW)[i] will contain x coordinate of i'th feature + //! keypoints.ptr(Y_ROW)[i] will contain y coordinate of i'th feature + //! keypoints.ptr(LAPLACIAN_ROW)[i] will contain laplacian sign of i'th feature + //! keypoints.ptr(OCTAVE_ROW)[i] will contain octave of i'th feature + //! keypoints.ptr(SIZE_ROW)[i] will contain size of i'th feature + //! keypoints.ptr(ANGLE_ROW)[i] will contain orientation of i'th feature + //! keypoints.ptr(HESSIAN_ROW)[i] will contain response of i'th feature + void operator()(const GpuMat& img, const GpuMat& mask, GpuMat& keypoints); + //! finds the keypoints and computes their descriptors. + //! Optionally it can compute descriptors for the user-provided keypoints and recompute keypoints direction + void operator()(const GpuMat& img, const GpuMat& mask, GpuMat& keypoints, GpuMat& descriptors, + bool useProvidedKeypoints = false); + + void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints); + void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints, GpuMat& descriptors, + bool useProvidedKeypoints = false); + + void operator()(const GpuMat& img, const GpuMat& mask, std::vector& keypoints, std::vector& descriptors, + bool useProvidedKeypoints = false); + + void releaseMemory(); + + // SURF parameters + double hessianThreshold; + int nOctaves; + int nOctaveLayers; + bool extended; + bool upright; + + //! max keypoints = min(keypointsRatio * img.size().area(), 65535) + float keypointsRatio; + + GpuMat sum, mask1, maskSum, intBuffer; + + GpuMat det, trace; + + GpuMat maxPosBuffer; +}; + +/*! + * The class implements the following algorithm: + * "ViBe: A universal background subtraction algorithm for video sequences" + * O. Barnich and M. Van D Roogenbroeck + * IEEE Transactions on Image Processing, 20(6) :1709-1724, June 2011 + */ +class CV_EXPORTS VIBE_GPU +{ +public: + //! the default constructor + explicit VIBE_GPU(unsigned long rngSeed = 1234567); + + //! re-initiaization method + void initialize(const GpuMat& firstFrame, Stream& stream = Stream::Null()); + + //! the update operator + void operator()(const GpuMat& frame, GpuMat& fgmask, Stream& stream = Stream::Null()); + + //! releases all inner buffers + void release(); + + int nbSamples; // number of samples per pixel + int reqMatches; // #_min + int radius; // R + int subsamplingFactor; // amount of random subsampling + +private: + Size frameSize_; + + unsigned long rngSeed_; + GpuMat randStates_; + + GpuMat samples_; +}; + +} // namespace gpu + +} // namespace cv + +#endif /* __OPENCV_GPUNONFREE_HPP__ */ diff --git a/modules/gpunonfree/perf/perf_gpunonfree.cpp b/modules/gpunonfree/perf/perf_gpunonfree.cpp new file mode 100644 index 0000000..4a27b77 --- /dev/null +++ b/modules/gpunonfree/perf/perf_gpunonfree.cpp @@ -0,0 +1,118 @@ +#include "perf_precomp.hpp" + +using namespace std; +using namespace testing; +using namespace perf; + +////////////////////////////////////////////////////////////////////// +// SURF + +DEF_PARAM_TEST_1(Image, string); + +PERF_TEST_P(Image, Features2D_SURF, + Values("gpu/perf/aloe.png")) +{ + declare.time(50.0); + + const cv::Mat img = readImage(GetParam(), cv::IMREAD_GRAYSCALE); + ASSERT_FALSE(img.empty()); + + if (PERF_RUN_GPU()) + { + cv::gpu::SURF_GPU d_surf; + + const cv::gpu::GpuMat d_img(img); + cv::gpu::GpuMat d_keypoints, d_descriptors; + + TEST_CYCLE() d_surf(d_img, cv::gpu::GpuMat(), d_keypoints, d_descriptors); + + std::vector gpu_keypoints; + d_surf.downloadKeypoints(d_keypoints, gpu_keypoints); + + cv::Mat gpu_descriptors(d_descriptors); + + sortKeyPoints(gpu_keypoints, gpu_descriptors); + + SANITY_CHECK_KEYPOINTS(gpu_keypoints); + SANITY_CHECK(gpu_descriptors, 1e-3); + } + else + { + cv::SURF surf; + + std::vector cpu_keypoints; + cv::Mat cpu_descriptors; + + TEST_CYCLE() surf(img, cv::noArray(), cpu_keypoints, cpu_descriptors); + + SANITY_CHECK_KEYPOINTS(cpu_keypoints); + SANITY_CHECK(cpu_descriptors); + } +} + +////////////////////////////////////////////////////// +// VIBE + +DEF_PARAM_TEST(Video_Cn, string, int); + +PERF_TEST_P(Video_Cn, Video_VIBE, + Combine(Values("gpu/video/768x576.avi", "gpu/video/1920x1080.avi"), + GPU_CHANNELS_1_3_4)) +{ + const string inputFile = perf::TestBase::getDataPath(GET_PARAM(0)); + const int cn = GET_PARAM(1); + + cv::VideoCapture cap(inputFile); + ASSERT_TRUE(cap.isOpened()); + + cv::Mat frame; + cap >> frame; + ASSERT_FALSE(frame.empty()); + + if (cn != 3) + { + cv::Mat temp; + if (cn == 1) + cv::cvtColor(frame, temp, cv::COLOR_BGR2GRAY); + else + cv::cvtColor(frame, temp, cv::COLOR_BGR2BGRA); + cv::swap(temp, frame); + } + + if (PERF_RUN_GPU()) + { + cv::gpu::GpuMat d_frame(frame); + cv::gpu::VIBE_GPU vibe; + cv::gpu::GpuMat foreground; + + vibe(d_frame, foreground); + + for (int i = 0; i < 10; ++i) + { + cap >> frame; + ASSERT_FALSE(frame.empty()); + + if (cn != 3) + { + cv::Mat temp; + if (cn == 1) + cv::cvtColor(frame, temp, cv::COLOR_BGR2GRAY); + else + cv::cvtColor(frame, temp, cv::COLOR_BGR2BGRA); + cv::swap(temp, frame); + } + + d_frame.upload(frame); + + startTimer(); next(); + vibe(d_frame, foreground); + stopTimer(); + } + + GPU_SANITY_CHECK(foreground); + } + else + { + FAIL_NO_CPU(); + } +} diff --git a/modules/gpunonfree/perf/perf_main.cpp b/modules/gpunonfree/perf/perf_main.cpp new file mode 100644 index 0000000..75e3ca6 --- /dev/null +++ b/modules/gpunonfree/perf/perf_main.cpp @@ -0,0 +1,5 @@ +#include "perf_precomp.hpp" + +using namespace perf; + +CV_PERF_TEST_MAIN(gpunonfree, printCudaInfo()) diff --git a/modules/gpunonfree/perf/perf_precomp.cpp b/modules/gpunonfree/perf/perf_precomp.cpp new file mode 100644 index 0000000..8552ac3 --- /dev/null +++ b/modules/gpunonfree/perf/perf_precomp.cpp @@ -0,0 +1 @@ +#include "perf_precomp.hpp" diff --git a/modules/gpunonfree/perf/perf_precomp.hpp b/modules/gpunonfree/perf/perf_precomp.hpp new file mode 100644 index 0000000..410eaec --- /dev/null +++ b/modules/gpunonfree/perf/perf_precomp.hpp @@ -0,0 +1,31 @@ +#ifdef __GNUC__ +# pragma GCC diagnostic ignored "-Wmissing-declarations" +# if defined __clang__ || defined __APPLE__ +# pragma GCC diagnostic ignored "-Wmissing-prototypes" +# pragma GCC diagnostic ignored "-Wextra" +# endif +#endif + +#ifndef __OPENCV_PERF_PRECOMP_HPP__ +#define __OPENCV_PERF_PRECOMP_HPP__ + +#include +#include + +#include "cvconfig.h" + +#include "opencv2/ts/ts.hpp" +#include "opencv2/ts/ts_perf.hpp" +#include "opencv2/ts/gpu_perf.hpp" + +#include "opencv2/core/core.hpp" +#include "opencv2/highgui/highgui.hpp" +#include "opencv2/gpu/gpu.hpp" +#include "opencv2/gpunonfree/gpunonfree.hpp" +#include "opencv2/nonfree/nonfree.hpp" + +#ifdef GTEST_CREATE_SHARED_LIBRARY +#error no modules except ts should have GTEST_CREATE_SHARED_LIBRARY defined +#endif + +#endif diff --git a/modules/gpu/src/cuda/surf.cu b/modules/gpunonfree/src/cuda/surf.cu similarity index 96% rename from modules/gpu/src/cuda/surf.cu rename to modules/gpunonfree/src/cuda/surf.cu index 5dc2c82..b82e2c3 100644 --- a/modules/gpu/src/cuda/surf.cu +++ b/modules/gpunonfree/src/cuda/surf.cu @@ -59,6 +59,33 @@ namespace cv { namespace gpu { namespace device { namespace surf { + void loadGlobalConstants(int maxCandidates, int maxFeatures, int img_rows, int img_cols, int nOctaveLayers, float hessianThreshold); + void loadOctaveConstants(int octave, int layer_rows, int layer_cols); + + void bindImgTex(PtrStepSzb img); + size_t bindSumTex(PtrStepSz sum); + size_t bindMaskSumTex(PtrStepSz maskSum); + + void icvCalcLayerDetAndTrace_gpu(const PtrStepf& det, const PtrStepf& trace, int img_rows, int img_cols, + int octave, int nOctaveLayer); + + void icvFindMaximaInLayer_gpu(const PtrStepf& det, const PtrStepf& trace, int4* maxPosBuffer, unsigned int* maxCounter, + int img_rows, int img_cols, int octave, bool use_mask, int nLayers); + + void icvInterpolateKeypoint_gpu(const PtrStepf& det, const int4* maxPosBuffer, unsigned int maxCounter, + float* featureX, float* featureY, int* featureLaplacian, int* featureOctave, float* featureSize, float* featureHessian, + unsigned int* featureCounter); + + void icvCalcOrientation_gpu(const float* featureX, const float* featureY, const float* featureSize, float* featureDir, int nFeatures); + + void compute_descriptors_gpu(PtrStepSz descriptors, const float* featureX, const float* featureY, const float* featureSize, const float* featureDir, int nFeatures); + } +}}} + +namespace cv { namespace gpu { namespace device +{ + namespace surf + { //////////////////////////////////////////////////////////////////////// // Global parameters @@ -718,6 +745,9 @@ namespace cv { namespace gpu { namespace device }; __device__ void calc_dx_dy(const float* featureX, const float* featureY, const float* featureSize, const float* featureDir, + float& dx, float& dy); + + __device__ void calc_dx_dy(const float* featureX, const float* featureY, const float* featureSize, const float* featureDir, float& dx, float& dy) { __shared__ float s_PATCH[PATCH_SZ + 1][PATCH_SZ + 1]; diff --git a/modules/gpu/src/cuda/bgfg_vibe.cu b/modules/gpunonfree/src/cuda/vibe.cu similarity index 95% rename from modules/gpu/src/cuda/bgfg_vibe.cu rename to modules/gpunonfree/src/cuda/vibe.cu index 72ea7f0..ff489ee 100644 --- a/modules/gpu/src/cuda/bgfg_vibe.cu +++ b/modules/gpunonfree/src/cuda/vibe.cu @@ -48,6 +48,18 @@ namespace cv { namespace gpu { namespace device { namespace vibe { + void loadConstants(int nbSamples, int reqMatches, int radius, int subsamplingFactor); + + void init_gpu(PtrStepSzb frame, int cn, PtrStepSzb samples, PtrStepSz randStates, cudaStream_t stream); + + void update_gpu(PtrStepSzb frame, int cn, PtrStepSzb fgmask, PtrStepSzb samples, PtrStepSz randStates, cudaStream_t stream); + } +}}} + +namespace cv { namespace gpu { namespace device +{ + namespace vibe + { __constant__ int c_nbSamples; __constant__ int c_reqMatches; __constant__ int c_radius; @@ -255,4 +267,4 @@ namespace cv { namespace gpu { namespace device }}} -#endif /* CUDA_DISABLER */ \ No newline at end of file +#endif /* CUDA_DISABLER */ diff --git a/modules/gpunonfree/src/precomp.cpp b/modules/gpunonfree/src/precomp.cpp new file mode 100644 index 0000000..75eb693 --- /dev/null +++ b/modules/gpunonfree/src/precomp.cpp @@ -0,0 +1,45 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// Intel License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000, Intel Corporation, all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials provided with the distribution. +// +// * The name of Intel Corporation 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. +// +//M*/ + +#include "precomp.hpp" + +/* End of file. */ + diff --git a/modules/gpunonfree/src/precomp.hpp b/modules/gpunonfree/src/precomp.hpp new file mode 100644 index 0000000..59a5a2c --- /dev/null +++ b/modules/gpunonfree/src/precomp.hpp @@ -0,0 +1,71 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000-2008, Intel Corporation, all rights reserved. +// Copyright (C) 2009, Willow Garage Inc., all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials 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. +// +//M*/ + +#ifndef __OPENCV_PRECOMP_H__ +#define __OPENCV_PRECOMP_H__ + +#if defined _MSC_VER && _MSC_VER >= 1200 + #pragma warning( disable: 4251 4710 4711 4514 4996 ) +#endif + +#ifdef HAVE_CVCONFIG_H + #include "cvconfig.h" +#endif + +#include + +#include "opencv2/gpu/gpu.hpp" +#include "opencv2/gpunonfree/gpunonfree.hpp" + +#ifdef HAVE_CUDA + #include "opencv2/gpu/stream_accessor.hpp" + #include "opencv2/gpu/device/common.hpp" + + static inline void throw_nogpu() { CV_Error(CV_StsNotImplemented, "The called functionality is disabled for current build or platform"); } + +#else /* defined(HAVE_CUDA) */ + + static inline void throw_nogpu() { CV_Error(CV_GpuNotSupported, "The library is compiled without GPU support"); } + +#endif /* defined(HAVE_CUDA) */ + +#endif /* __OPENCV_PRECOMP_H__ */ diff --git a/modules/gpu/src/surf.cpp b/modules/gpunonfree/src/surf.cpp similarity index 100% rename from modules/gpu/src/surf.cpp rename to modules/gpunonfree/src/surf.cpp diff --git a/modules/gpu/src/bgfg_vibe.cpp b/modules/gpunonfree/src/vibe.cpp similarity index 100% rename from modules/gpu/src/bgfg_vibe.cpp rename to modules/gpunonfree/src/vibe.cpp diff --git a/modules/gpunonfree/test/test_gpunonfree.cpp b/modules/gpunonfree/test/test_gpunonfree.cpp new file mode 100644 index 0000000..2d4e179 --- /dev/null +++ b/modules/gpunonfree/test/test_gpunonfree.cpp @@ -0,0 +1,286 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// Intel License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000, Intel Corporation, all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials provided with the distribution. +// +// * The name of Intel Corporation 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. +// +//M*/ + +#include "test_precomp.hpp" + +#ifdef HAVE_CUDA + +using namespace cvtest; + +///////////////////////////////////////////////////////////////////////////////////////////////// +// SURF + +namespace +{ + IMPLEMENT_PARAM_CLASS(SURF_HessianThreshold, double) + IMPLEMENT_PARAM_CLASS(SURF_Octaves, int) + IMPLEMENT_PARAM_CLASS(SURF_OctaveLayers, int) + IMPLEMENT_PARAM_CLASS(SURF_Extended, bool) + IMPLEMENT_PARAM_CLASS(SURF_Upright, bool) +} + +PARAM_TEST_CASE(SURF, cv::gpu::DeviceInfo, SURF_HessianThreshold, SURF_Octaves, SURF_OctaveLayers, SURF_Extended, SURF_Upright) +{ + cv::gpu::DeviceInfo devInfo; + double hessianThreshold; + int nOctaves; + int nOctaveLayers; + bool extended; + bool upright; + + virtual void SetUp() + { + devInfo = GET_PARAM(0); + hessianThreshold = GET_PARAM(1); + nOctaves = GET_PARAM(2); + nOctaveLayers = GET_PARAM(3); + extended = GET_PARAM(4); + upright = GET_PARAM(5); + + cv::gpu::setDevice(devInfo.deviceID()); + } +}; + +GPU_TEST_P(SURF, Detector) +{ + cv::Mat image = readImage("features2d/aloe.png", cv::IMREAD_GRAYSCALE); + ASSERT_FALSE(image.empty()); + + cv::gpu::SURF_GPU surf; + surf.hessianThreshold = hessianThreshold; + surf.nOctaves = nOctaves; + surf.nOctaveLayers = nOctaveLayers; + surf.extended = extended; + surf.upright = upright; + surf.keypointsRatio = 0.05f; + + if (!supportFeature(devInfo, cv::gpu::GLOBAL_ATOMICS)) + { + try + { + std::vector keypoints; + surf(loadMat(image), cv::gpu::GpuMat(), keypoints); + } + catch (const cv::Exception& e) + { + ASSERT_EQ(CV_StsNotImplemented, e.code); + } + } + else + { + std::vector keypoints; + surf(loadMat(image), cv::gpu::GpuMat(), keypoints); + + cv::SURF surf_gold; + surf_gold.hessianThreshold = hessianThreshold; + surf_gold.nOctaves = nOctaves; + surf_gold.nOctaveLayers = nOctaveLayers; + surf_gold.extended = extended; + surf_gold.upright = upright; + + std::vector keypoints_gold; + surf_gold(image, cv::noArray(), keypoints_gold); + + ASSERT_EQ(keypoints_gold.size(), keypoints.size()); + int matchedCount = getMatchedPointsCount(keypoints_gold, keypoints); + double matchedRatio = static_cast(matchedCount) / keypoints_gold.size(); + + EXPECT_GT(matchedRatio, 0.95); + } +} + +GPU_TEST_P(SURF, Detector_Masked) +{ + cv::Mat image = readImage("features2d/aloe.png", cv::IMREAD_GRAYSCALE); + ASSERT_FALSE(image.empty()); + + cv::Mat mask(image.size(), CV_8UC1, cv::Scalar::all(1)); + mask(cv::Range(0, image.rows / 2), cv::Range(0, image.cols / 2)).setTo(cv::Scalar::all(0)); + + cv::gpu::SURF_GPU surf; + surf.hessianThreshold = hessianThreshold; + surf.nOctaves = nOctaves; + surf.nOctaveLayers = nOctaveLayers; + surf.extended = extended; + surf.upright = upright; + surf.keypointsRatio = 0.05f; + + if (!supportFeature(devInfo, cv::gpu::GLOBAL_ATOMICS)) + { + try + { + std::vector keypoints; + surf(loadMat(image), loadMat(mask), keypoints); + } + catch (const cv::Exception& e) + { + ASSERT_EQ(CV_StsNotImplemented, e.code); + } + } + else + { + std::vector keypoints; + surf(loadMat(image), loadMat(mask), keypoints); + + cv::SURF surf_gold; + surf_gold.hessianThreshold = hessianThreshold; + surf_gold.nOctaves = nOctaves; + surf_gold.nOctaveLayers = nOctaveLayers; + surf_gold.extended = extended; + surf_gold.upright = upright; + + std::vector keypoints_gold; + surf_gold(image, mask, keypoints_gold); + + ASSERT_EQ(keypoints_gold.size(), keypoints.size()); + int matchedCount = getMatchedPointsCount(keypoints_gold, keypoints); + double matchedRatio = static_cast(matchedCount) / keypoints_gold.size(); + + EXPECT_GT(matchedRatio, 0.95); + } +} + +GPU_TEST_P(SURF, Descriptor) +{ + cv::Mat image = readImage("features2d/aloe.png", cv::IMREAD_GRAYSCALE); + ASSERT_FALSE(image.empty()); + + cv::gpu::SURF_GPU surf; + surf.hessianThreshold = hessianThreshold; + surf.nOctaves = nOctaves; + surf.nOctaveLayers = nOctaveLayers; + surf.extended = extended; + surf.upright = upright; + surf.keypointsRatio = 0.05f; + + cv::SURF surf_gold; + surf_gold.hessianThreshold = hessianThreshold; + surf_gold.nOctaves = nOctaves; + surf_gold.nOctaveLayers = nOctaveLayers; + surf_gold.extended = extended; + surf_gold.upright = upright; + + if (!supportFeature(devInfo, cv::gpu::GLOBAL_ATOMICS)) + { + try + { + std::vector keypoints; + cv::gpu::GpuMat descriptors; + surf(loadMat(image), cv::gpu::GpuMat(), keypoints, descriptors); + } + catch (const cv::Exception& e) + { + ASSERT_EQ(CV_StsNotImplemented, e.code); + } + } + else + { + std::vector keypoints; + surf_gold(image, cv::noArray(), keypoints); + + cv::gpu::GpuMat descriptors; + surf(loadMat(image), cv::gpu::GpuMat(), keypoints, descriptors, true); + + cv::Mat descriptors_gold; + surf_gold(image, cv::noArray(), keypoints, descriptors_gold, true); + + cv::BFMatcher matcher(cv::NORM_L2); + std::vector matches; + matcher.match(descriptors_gold, cv::Mat(descriptors), matches); + + int matchedCount = getMatchedPointsCount(keypoints, keypoints, matches); + double matchedRatio = static_cast(matchedCount) / keypoints.size(); + + EXPECT_GT(matchedRatio, 0.6); + } +} + +INSTANTIATE_TEST_CASE_P(GPU_Features2D, SURF, testing::Combine( + ALL_DEVICES, + testing::Values(SURF_HessianThreshold(100.0), SURF_HessianThreshold(500.0), SURF_HessianThreshold(1000.0)), + testing::Values(SURF_Octaves(3), SURF_Octaves(4)), + testing::Values(SURF_OctaveLayers(2), SURF_OctaveLayers(3)), + testing::Values(SURF_Extended(false), SURF_Extended(true)), + testing::Values(SURF_Upright(false), SURF_Upright(true)))); + +////////////////////////////////////////////////////// +// VIBE + +PARAM_TEST_CASE(VIBE, cv::gpu::DeviceInfo, cv::Size, MatType, UseRoi) +{ +}; + +GPU_TEST_P(VIBE, Accuracy) +{ + const cv::gpu::DeviceInfo devInfo = GET_PARAM(0); + cv::gpu::setDevice(devInfo.deviceID()); + const cv::Size size = GET_PARAM(1); + const int type = GET_PARAM(2); + const bool useRoi = GET_PARAM(3); + + const cv::Mat fullfg(size, CV_8UC1, cv::Scalar::all(255)); + + cv::Mat frame = randomMat(size, type, 0.0, 100); + cv::gpu::GpuMat d_frame = loadMat(frame, useRoi); + + cv::gpu::VIBE_GPU vibe; + cv::gpu::GpuMat d_fgmask = createMat(size, CV_8UC1, useRoi); + vibe.initialize(d_frame); + + for (int i = 0; i < 20; ++i) + vibe(d_frame, d_fgmask); + + frame = randomMat(size, type, 160, 255); + d_frame = loadMat(frame, useRoi); + vibe(d_frame, d_fgmask); + + // now fgmask should be entirely foreground + ASSERT_MAT_NEAR(fullfg, d_fgmask, 0); +} + +INSTANTIATE_TEST_CASE_P(GPU_Video, VIBE, testing::Combine( + ALL_DEVICES, + DIFFERENT_SIZES, + testing::Values(MatType(CV_8UC1), MatType(CV_8UC3), MatType(CV_8UC4)), + WHOLE_SUBMAT)); + +#endif // HAVE_CUDA + diff --git a/modules/gpunonfree/test/test_main.cpp b/modules/gpunonfree/test/test_main.cpp new file mode 100644 index 0000000..4c15ff9 --- /dev/null +++ b/modules/gpunonfree/test/test_main.cpp @@ -0,0 +1,119 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// Intel License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000, Intel Corporation, all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials provided with the distribution. +// +// * The name of Intel Corporation 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. +// +//M*/ + +#include "test_precomp.hpp" + +#ifdef HAVE_CUDA + +using namespace std; +using namespace cv; +using namespace cv::gpu; +using namespace cvtest; +using namespace testing; + +int main(int argc, char** argv) +{ + try + { + const char* keys = + "{ h | help ? | false | Print help}" + "{ i | info | false | Print information about system and exit }" + "{ d | device | -1 | Device on which tests will be executed (-1 means all devices) }" + ; + + CommandLineParser cmd(argc, (const char**)argv, keys); + + if (cmd.get("help")) + { + cmd.printParams(); + return 0; + } + + printCudaInfo(); + + if (cmd.get("info")) + { + return 0; + } + + int device = cmd.get("device"); + if (device < 0) + { + DeviceManager::instance().loadAll(); + + cout << "Run tests on all supported devices \n" << endl; + } + else + { + DeviceManager::instance().load(device); + + DeviceInfo info(device); + cout << "Run tests on device " << device << " [" << info.name() << "] \n" << endl; + } + + TS::ptr()->init("gpu"); + InitGoogleTest(&argc, argv); + + return RUN_ALL_TESTS(); + } + catch (const exception& e) + { + cerr << e.what() << endl; + return -1; + } + catch (...) + { + cerr << "Unknown error" << endl; + return -1; + } + + return 0; +} + +#else // HAVE_CUDA + +int main() +{ + std::cerr << "OpenCV was built without CUDA support" << std::endl; + return 0; +} + +#endif // HAVE_CUDA diff --git a/modules/gpunonfree/test/test_precomp.cpp b/modules/gpunonfree/test/test_precomp.cpp new file mode 100644 index 0000000..8121ae4 --- /dev/null +++ b/modules/gpunonfree/test/test_precomp.cpp @@ -0,0 +1,42 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// Intel License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000, Intel Corporation, all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials provided with the distribution. +// +// * The name of Intel Corporation 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. +// +//M*/ + +#include "test_precomp.hpp" diff --git a/modules/gpunonfree/test/test_precomp.hpp b/modules/gpunonfree/test/test_precomp.hpp new file mode 100644 index 0000000..5030a0d --- /dev/null +++ b/modules/gpunonfree/test/test_precomp.hpp @@ -0,0 +1,69 @@ +/*M/////////////////////////////////////////////////////////////////////////////////////// +// +// IMPORTANT: READ BEFORE DOWNLOADING, COPYING, INSTALLING OR USING. +// +// By downloading, copying, installing or using the software you agree to this license. +// If you do not agree to this license, do not download, install, +// copy or use the software. +// +// +// Intel License Agreement +// For Open Source Computer Vision Library +// +// Copyright (C) 2000, Intel Corporation, all rights reserved. +// Third party copyrights are property of their respective owners. +// +// 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 materials provided with the distribution. +// +// * The name of Intel Corporation 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. +// +//M*/ + +#ifdef __GNUC__ +# pragma GCC diagnostic ignored "-Wmissing-declarations" +# if defined __clang__ || defined __APPLE__ +# pragma GCC diagnostic ignored "-Wmissing-prototypes" +# pragma GCC diagnostic ignored "-Wextra" +# endif +#endif + +#ifndef __OPENCV_TEST_PRECOMP_HPP__ +#define __OPENCV_TEST_PRECOMP_HPP__ + +#include "cvconfig.h" + +#include + +#ifdef HAVE_CUDA + #include "opencv2/ts/ts.hpp" + #include "opencv2/ts/ts_perf.hpp" + #include "opencv2/ts/gpu_test.hpp" + + #include "opencv2/core/core.hpp" + #include "opencv2/highgui/highgui.hpp" + #include "opencv2/gpu/gpu.hpp" + #include "opencv2/gpunonfree/gpunonfree.hpp" + #include "opencv2/nonfree/nonfree.hpp" +#endif + +#endif diff --git a/modules/stitching/CMakeLists.txt b/modules/stitching/CMakeLists.txt index 5d48bd3..b310a8e 100644 --- a/modules/stitching/CMakeLists.txt +++ b/modules/stitching/CMakeLists.txt @@ -1,3 +1,3 @@ set(the_description "Images stitching") -ocv_define_module(stitching opencv_imgproc opencv_features2d opencv_calib3d opencv_objdetect OPTIONAL opencv_gpu opencv_nonfree) +ocv_define_module(stitching opencv_imgproc opencv_features2d opencv_calib3d opencv_objdetect OPTIONAL opencv_gpu opencv_gpunonfree opencv_nonfree) diff --git a/modules/stitching/include/opencv2/stitching/detail/matchers.hpp b/modules/stitching/include/opencv2/stitching/detail/matchers.hpp index f9a53ef..9332d94 100644 --- a/modules/stitching/include/opencv2/stitching/detail/matchers.hpp +++ b/modules/stitching/include/opencv2/stitching/detail/matchers.hpp @@ -47,8 +47,8 @@ #include "opencv2/features2d/features2d.hpp" #include "opencv2/opencv_modules.hpp" -#ifdef HAVE_OPENCV_GPU -#include "opencv2/gpu/gpu.hpp" +#ifdef HAVE_OPENCV_GPUNONFREE +#include "opencv2/gpunonfree/gpunonfree.hpp" #endif namespace cv { @@ -103,7 +103,7 @@ private: }; -#ifdef HAVE_OPENCV_GPU +#ifdef HAVE_OPENCV_GPUNONFREE class CV_EXPORTS SurfFeaturesFinderGpu : public FeaturesFinder { public: diff --git a/modules/stitching/src/matchers.cpp b/modules/stitching/src/matchers.cpp index dc6b5fa..07845d2 100644 --- a/modules/stitching/src/matchers.cpp +++ b/modules/stitching/src/matchers.cpp @@ -429,7 +429,7 @@ void OrbFeaturesFinder::find(const Mat &image, ImageFeatures &features) } } -#ifdef HAVE_OPENCV_GPU +#ifdef HAVE_OPENCV_GPUNONFREE SurfFeaturesFinderGpu::SurfFeaturesFinderGpu(double hess_thresh, int num_octaves, int num_layers, int num_octaves_descr, int num_layers_descr) { diff --git a/modules/stitching/src/precomp.hpp b/modules/stitching/src/precomp.hpp index 80eb186..0aade9f 100644 --- a/modules/stitching/src/precomp.hpp +++ b/modules/stitching/src/precomp.hpp @@ -73,6 +73,9 @@ #ifdef HAVE_OPENCV_GPU # include "opencv2/gpu/gpu.hpp" #endif +#ifdef HAVE_OPENCV_GPUNONFREE +# include "opencv2/gpunonfree/gpunonfree.hpp" +#endif #include "../../imgproc/src/gcgraph.hpp" diff --git a/modules/stitching/src/stitcher.cpp b/modules/stitching/src/stitcher.cpp index 2b9573b..2a2fe0f 100644 --- a/modules/stitching/src/stitcher.cpp +++ b/modules/stitching/src/stitcher.cpp @@ -61,7 +61,7 @@ Stitcher Stitcher::createDefault(bool try_use_gpu) #ifdef HAVE_OPENCV_GPU if (try_use_gpu && gpu::getCudaEnabledDeviceCount() > 0) { -#ifdef HAVE_OPENCV_NONFREE +#ifdef HAVE_OPENCV_GPUNONFREE stitcher.setFeaturesFinder(new detail::SurfFeaturesFinderGpu()); #else stitcher.setFeaturesFinder(new detail::OrbFeaturesFinder()); diff --git a/samples/cpp/CMakeLists.txt b/samples/cpp/CMakeLists.txt index 2004857..3f42f43 100644 --- a/samples/cpp/CMakeLists.txt +++ b/samples/cpp/CMakeLists.txt @@ -16,10 +16,14 @@ if(BUILD_EXAMPLES AND OCV_DEPENDENCIES_FOUND) ocv_include_directories("${OpenCV_SOURCE_DIR}/include")#for opencv.hpp ocv_include_modules(${OPENCV_CPP_SAMPLES_REQUIRED_DEPS}) - if (HAVE_opencv_gpu) + if(HAVE_opencv_gpu) ocv_include_directories("${OpenCV_SOURCE_DIR}/modules/gpu/include") endif() + if(HAVE_opencv_gpunonfree) + ocv_include_directories("${OpenCV_SOURCE_DIR}/modules/gpunonfree/include") + endif() + if(CMAKE_COMPILER_IS_GNUCXX AND NOT ENABLE_NOISY_WARNINGS) set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} -Wno-unused-function") endif() @@ -41,9 +45,12 @@ if(BUILD_EXAMPLES AND OCV_DEPENDENCIES_FOUND) add_executable(${the_target} ${srcs}) target_link_libraries(${the_target} ${OPENCV_LINKER_LIBS} ${OPENCV_CPP_SAMPLES_REQUIRED_DEPS}) - if (HAVE_opencv_gpu) + if(HAVE_opencv_gpu) target_link_libraries(${the_target} opencv_gpu) endif() + if(HAVE_opencv_gpunonfree) + target_link_libraries(${the_target} opencv_gpunonfree) + endif() set_target_properties(${the_target} PROPERTIES OUTPUT_NAME "cpp-${sample_kind}-${name}" diff --git a/samples/cpp/stitching_detailed.cpp b/samples/cpp/stitching_detailed.cpp index 73d139e..fe4ad87 100644 --- a/samples/cpp/stitching_detailed.cpp +++ b/samples/cpp/stitching_detailed.cpp @@ -355,7 +355,7 @@ int main(int argc, char* argv[]) Ptr finder; if (features_type == "surf") { -#ifdef HAVE_OPENCV_GPU +#ifdef HAVE_OPENCV_GPUNONFREE if (try_gpu && gpu::getCudaEnabledDeviceCount() > 0) finder = new SurfFeaturesFinderGpu(); else diff --git a/samples/gpu/CMakeLists.txt b/samples/gpu/CMakeLists.txt index 9fd19e2..f68d2cd 100644 --- a/samples/gpu/CMakeLists.txt +++ b/samples/gpu/CMakeLists.txt @@ -1,7 +1,7 @@ SET(OPENCV_GPU_SAMPLES_REQUIRED_DEPS opencv_core opencv_flann opencv_imgproc opencv_highgui opencv_ml opencv_video opencv_objdetect opencv_features2d opencv_calib3d opencv_legacy opencv_contrib opencv_gpu - opencv_nonfree opencv_superres) + opencv_superres) ocv_check_dependencies(${OPENCV_GPU_SAMPLES_REQUIRED_DEPS}) @@ -17,6 +17,10 @@ if(BUILD_EXAMPLES AND OCV_DEPENDENCIES_FOUND) "${OpenCV_SOURCE_DIR}/modules/gpu/src/nvidia/core" ) + if(HAVE_opencv_gpunonfree) + ocv_include_directories("${OpenCV_SOURCE_DIR}/modules/gpunonfree/include") + endif() + if(HAVE_CUDA) ocv_include_directories(${CUDA_INCLUDE_DIRS}) endif() @@ -33,6 +37,9 @@ if(BUILD_EXAMPLES AND OCV_DEPENDENCIES_FOUND) add_executable(${the_target} ${srcs}) target_link_libraries(${the_target} ${OPENCV_LINKER_LIBS} ${OPENCV_GPU_SAMPLES_REQUIRED_DEPS}) + if(HAVE_opencv_gpunonfree) + target_link_libraries(${the_target} opencv_gpunonfree) + endif() set_target_properties(${the_target} PROPERTIES OUTPUT_NAME "${project}-example-${name}" diff --git a/samples/gpu/bgfg_segm.cpp b/samples/gpu/bgfg_segm.cpp index 4e4c520..ff538f8 100644 --- a/samples/gpu/bgfg_segm.cpp +++ b/samples/gpu/bgfg_segm.cpp @@ -1,10 +1,15 @@ #include #include +#include "opencv2/opencv_modules.hpp" #include "opencv2/core/core.hpp" #include "opencv2/gpu/gpu.hpp" #include "opencv2/highgui/highgui.hpp" +#ifdef HAVE_OPENCV_GPUNONFREE +#include "opencv2/gpunonfree/gpunonfree.hpp" +#endif + using namespace std; using namespace cv; using namespace cv::gpu; @@ -14,7 +19,9 @@ enum Method FGD_STAT, MOG, MOG2, +#ifdef HAVE_OPENCV_GPUNONFREE VIBE, +#endif GMG }; @@ -38,13 +45,25 @@ int main(int argc, const char** argv) string file = cmd.get("file"); string method = cmd.get("method"); - if (method != "fgd" && method != "mog" && method != "mog2" && method != "vibe" && method != "gmg") + if (method != "fgd" + && method != "mog" + && method != "mog2" + #ifdef HAVE_OPENCV_GPUNONFREE + && method != "vibe" + #endif + && method != "gmg") { cerr << "Incorrect method" << endl; return -1; } - Method m = method == "fgd" ? FGD_STAT : method == "mog" ? MOG : method == "mog2" ? MOG2 : method == "vibe" ? VIBE : GMG; + Method m = method == "fgd" ? FGD_STAT : + method == "mog" ? MOG : + method == "mog2" ? MOG2 : + #ifdef HAVE_OPENCV_GPUNONFREE + method == "vibe" ? VIBE : + #endif + GMG; VideoCapture cap; @@ -67,7 +86,9 @@ int main(int argc, const char** argv) FGDStatModel fgd_stat; MOG_GPU mog; MOG2_GPU mog2; +#ifdef HAVE_OPENCV_GPUNONFREE VIBE_GPU vibe; +#endif GMG_GPU gmg; gmg.numInitializationFrames = 40; @@ -93,9 +114,11 @@ int main(int argc, const char** argv) mog2(d_frame, d_fgmask); break; +#ifdef HAVE_OPENCV_GPUNONFREE case VIBE: vibe.initialize(d_frame); break; +#endif case GMG: gmg.initialize(d_frame.size()); @@ -105,8 +128,14 @@ int main(int argc, const char** argv) namedWindow("image", WINDOW_NORMAL); namedWindow("foreground mask", WINDOW_NORMAL); namedWindow("foreground image", WINDOW_NORMAL); - if (m != VIBE && m != GMG) + if (m != GMG + #ifdef HAVE_OPENCV_GPUNONFREE + && m != VIBE + #endif + ) + { namedWindow("mean background image", WINDOW_NORMAL); + } for(;;) { @@ -136,9 +165,11 @@ int main(int argc, const char** argv) mog2.getBackgroundImage(d_bgimg); break; +#ifdef HAVE_OPENCV_GPUNONFREE case VIBE: vibe(d_frame, d_fgmask); break; +#endif case GMG: gmg(d_frame, d_fgmask); diff --git a/samples/gpu/driver_api_stereo_multi.cpp b/samples/gpu/driver_api_stereo_multi.cpp index b8f99e8..10c3974 100644 --- a/samples/gpu/driver_api_stereo_multi.cpp +++ b/samples/gpu/driver_api_stereo_multi.cpp @@ -73,9 +73,6 @@ GpuMat d_right[2]; StereoBM_GPU* bm[2]; GpuMat d_result[2]; -// CPU result -Mat result; - static void printHelp() { std::cout << "Usage: driver_api_stereo_multi_gpu --left --right \n"; diff --git a/samples/gpu/performance/CMakeLists.txt b/samples/gpu/performance/CMakeLists.txt index 3b3d6e4..2e8677d 100644 --- a/samples/gpu/performance/CMakeLists.txt +++ b/samples/gpu/performance/CMakeLists.txt @@ -3,9 +3,17 @@ set(the_target "example_gpu_performance") file(GLOB sources "performance/*.cpp") file(GLOB headers "performance/*.h") +if(HAVE_opencv_nonfree) + ocv_include_directories("${OpenCV_SOURCE_DIR}/modules/nonfree/include") +endif() + add_executable(${the_target} ${sources} ${headers}) target_link_libraries(${the_target} ${OPENCV_LINKER_LIBS} ${OPENCV_GPU_SAMPLES_REQUIRED_DEPS}) +if(HAVE_opencv_gpunonfree) + target_link_libraries(${the_target} opencv_gpunonfree opencv_nonfree) +endif() + set_target_properties(${the_target} PROPERTIES OUTPUT_NAME "performance_gpu" PROJECT_LABEL "(EXAMPLE_GPU) performance") diff --git a/samples/gpu/performance/tests.cpp b/samples/gpu/performance/tests.cpp index b35102d..3c77d59 100644 --- a/samples/gpu/performance/tests.cpp +++ b/samples/gpu/performance/tests.cpp @@ -4,10 +4,15 @@ #include "opencv2/calib3d/calib3d.hpp" #include "opencv2/video/video.hpp" #include "opencv2/gpu/gpu.hpp" -#include "opencv2/nonfree/nonfree.hpp" #include "opencv2/legacy/legacy.hpp" #include "performance.h" +#include "opencv2/opencv_modules.hpp" +#ifdef HAVE_OPENCV_GPUNONFREE +#include "opencv2/gpunonfree/gpunonfree.hpp" +#include "opencv2/nonfree/nonfree.hpp" +#endif + using namespace std; using namespace cv; @@ -266,6 +271,7 @@ TEST(meanShift) } } +#ifdef HAVE_OPENCV_GPUNONFREE TEST(SURF) { @@ -294,6 +300,8 @@ TEST(SURF) GPU_OFF; } +#endif + TEST(FAST) { diff --git a/samples/gpu/stereo_multi.cpp b/samples/gpu/stereo_multi.cpp index d424bf9..cf42043 100644 --- a/samples/gpu/stereo_multi.cpp +++ b/samples/gpu/stereo_multi.cpp @@ -44,9 +44,6 @@ GpuMat d_right[2]; StereoBM_GPU* bm[2]; GpuMat d_result[2]; -// CPU result -Mat result; - static void printHelp() { std::cout << "Usage: stereo_multi_gpu --left --right \n"; diff --git a/samples/gpu/surf_keypoint_matcher.cpp b/samples/gpu/surf_keypoint_matcher.cpp index 617cda5..914867b 100644 --- a/samples/gpu/surf_keypoint_matcher.cpp +++ b/samples/gpu/surf_keypoint_matcher.cpp @@ -1,9 +1,14 @@ #include +#include "opencv2/opencv_modules.hpp" + +#ifdef HAVE_OPENCV_GPUNONFREE + #include "opencv2/core/core.hpp" #include "opencv2/features2d/features2d.hpp" #include "opencv2/highgui/highgui.hpp" #include "opencv2/gpu/gpu.hpp" +#include "opencv2/gpunonfree/gpunonfree.hpp" using namespace std; using namespace cv; @@ -81,3 +86,13 @@ int main(int argc, char* argv[]) return 0; } + +#else + +int main() +{ + std::cerr << "OpenCV was built without gpunonfree module" << std::endl; + return 0; +} + +#endif -- 2.7.4