added NVIDIA tests (disabled because doesn't work under Linux)

modules/gpu/test/CMakeLists.txt

@@ -19,20 +19,29 @@ endforeach()
 file(GLOB test_srcs "*.cpp")
 file(GLOB test_hdrs "*.h*")
 
-add_executable(${the_target} ${test_srcs} ${test_hdrs})
-
-if(PCHSupport_FOUND)
-    set(pch_header ${CMAKE_CURRENT_SOURCE_DIR}/test_precomp.hpp)
-    if(${CMAKE_GENERATOR} MATCHES "Visual*" OR ${CMAKE_GENERATOR} MATCHES "Xcode*")
-        if(${CMAKE_GENERATOR} MATCHES "Visual*")
-            set(${the_target}_pch "test_precomp.cpp")
-        endif()            
-        add_native_precompiled_header(${the_target} ${pch_header})
-    elseif(CMAKE_COMPILER_IS_GNUCXX AND ${CMAKE_GENERATOR} MATCHES ".*Makefiles")
-        add_precompiled_header(${the_target} ${pch_header})
-    endif()
+if(HAVE_CUDA)
+        include_directories(${CUDA_INCLUDE_DIRS} ${CMAKE_SOURCE_DIR}/modules/gpu/src/nvidia ${CMAKE_SOURCE_DIR}/modules/gpu/src/nvidia/core ${CMAKE_SOURCE_DIR}/modules/gpu/src/nvidia/NPP_staging)
+
+        file(GLOB nvidia "nvidia/*.*")
+        SET(ncv_cpp ${CMAKE_SOURCE_DIR}/modules/gpu/src/nvidia/core/NCV.cpp)
+        source_group("nvidia" FILES ${nvidia})
 endif()
 
+
+add_executable(${the_target} ${test_srcs} ${test_hdrs} ${nvidia} ${ncv_cpp})
+
+#if(PCHSupport_FOUND)
+#    set(pch_header ${CMAKE_CURRENT_SOURCE_DIR}/test_precomp.hpp)
+#    if(${CMAKE_GENERATOR} MATCHES "Visual*" OR ${CMAKE_GENERATOR} MATCHES "Xcode*")
+#        if(${CMAKE_GENERATOR} MATCHES "Visual*")
+#            set(${the_target}_pch "test_precomp.cpp")
+#        endif()            
+#        add_native_precompiled_header(${the_target} ${pch_header})
+#    elseif(CMAKE_COMPILER_IS_GNUCXX AND ${CMAKE_GENERATOR} MATCHES ".*Makefiles")
+#        add_precompiled_header(${the_target} ${pch_header})
+#    endif()
+#endif()
+
 # Additional target properties
 set_target_properties(${the_target} PROPERTIES
     DEBUG_POSTFIX "${OPENCV_DEBUG_POSTFIX}"

modules/gpu/test/nvidia/NCVAutoTestLister.hpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _ncvautotestlister_hpp_
+#define _ncvautotestlister_hpp_
+
+#include <vector>
+
+#include "NCVTest.hpp"
+
+class NCVAutoTestLister
+{
+public:
+
+    NCVAutoTestLister(std::string testSuiteName, NcvBool bStopOnFirstFail=false, NcvBool bCompactOutput=true)
+        :
+    testSuiteName(testSuiteName),
+    bStopOnFirstFail(bStopOnFirstFail),
+    bCompactOutput(bCompactOutput)
+    {
+    }
+
+    void add(INCVTest *test)
+    {
+        this->tests.push_back(test);
+    }
+
+    bool invoke()
+    {
+        Ncv32u nPassed = 0;
+        Ncv32u nFailed = 0;
+        Ncv32u nFailedMem = 0;
+
+        if (bCompactOutput)
+        {
+            printf("Test suite '%s' with %d tests\n", 
+                testSuiteName.c_str(),
+                (int)(this->tests.size()));
+        }
+
+        for (Ncv32u i=0; i<this->tests.size(); i++)
+        {
+            INCVTest &curTest = *tests[i];
+
+            NCVTestReport curReport;
+            bool res = curTest.executeTest(curReport);
+
+            if (!bCompactOutput)
+            {
+                printf("Test %3i %16s; Consumed mem GPU = %8d, CPU = %8d; %s\n",
+                    i,
+                    curTest.getName().c_str(),
+                    curReport.statsNums["MemGPU"],
+                    curReport.statsNums["MemCPU"],
+                    curReport.statsText["rcode"].c_str());
+            }
+
+            if (res)
+            {
+                nPassed++;
+                if (bCompactOutput)
+                {
+                    printf(".");
+                }
+            }
+            else
+            {
+                if (!curReport.statsText["rcode"].compare("FAILED"))
+                {
+                    nFailed++;
+                    if (bCompactOutput)
+                    {
+                        printf("x");
+                    }
+                    if (bStopOnFirstFail)
+                    {
+                        break;
+                    }
+                }
+                else
+                {
+                    nFailedMem++;
+                    if (bCompactOutput)
+                    {
+                        printf("m");
+                    }
+                }
+            }
+            fflush(stdout);
+        }
+        if (bCompactOutput)
+        {
+            printf("\n");
+        }
+
+        printf("Test suite '%s' complete: %d total, %d passed, %d memory errors, %d failed\n\n", 
+            testSuiteName.c_str(),
+            (int)(this->tests.size()),
+            nPassed,
+            nFailedMem,
+            nFailed);
+
+        bool passed = nFailed == 0 && nFailedMem == 0;
+        return passed;
+    }
+
+    ~NCVAutoTestLister()
+    {
+        for (Ncv32u i=0; i<this->tests.size(); i++)
+        {
+            delete tests[i];
+        }
+    }
+
+private:
+
+    NcvBool bStopOnFirstFail;
+    NcvBool bCompactOutput;
+    std::string testSuiteName;
+    std::vector<INCVTest *> tests;
+};
+
+#endif // _ncvautotestlister_hpp_

modules/gpu/test/nvidia/NCVTest.hpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _ncvtest_hpp_
+#define _ncvtest_hpp_
+
+#pragma warning( disable : 4201 4408 4127 4100)
+
+#include <string>
+#include <vector>
+#include <map>
+#include <memory>
+#include <algorithm>
+#include <fstream>
+
+#include <cuda_runtime.h>
+#include "NPP_staging.hpp"
+
+
+struct NCVTestReport
+{
+    std::map<std::string, Ncv32u> statsNums;
+    std::map<std::string, std::string> statsText;
+};
+
+
+class INCVTest
+{
+public:
+    virtual bool executeTest(NCVTestReport &report) = 0;
+    virtual std::string getName() const = 0;
+};
+
+
+class NCVTestProvider : public INCVTest
+{
+public:
+
+    NCVTestProvider(std::string testName)
+        :
+        testName(testName)
+    {
+        int devId;
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDevice(&devId), "Error returned from cudaGetDevice", );
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDeviceProperties(&this->devProp, devId), "Error returned from cudaGetDeviceProperties", );
+    }
+
+    virtual bool init() = 0;
+    virtual bool process() = 0;
+    virtual bool deinit() = 0;
+    virtual bool toString(std::ofstream &strOut) = 0;
+
+    virtual std::string getName() const
+    {
+        return this->testName;
+    }
+
+    virtual ~NCVTestProvider()
+    {
+        deinitMemory();
+    }
+
+    virtual bool executeTest(NCVTestReport &report)
+    {
+        bool res;
+        report.statsText["rcode"] = "FAILED";
+
+        res = initMemory(report);
+        if (!res)
+        {
+            dumpToFile(report);
+            deinitMemory();
+            return false;
+        }
+
+        res = init();
+        if (!res)
+        {
+            dumpToFile(report);
+            deinit();
+            deinitMemory();
+            return false;
+        }
+
+        res = process();
+        if (!res)
+        {
+            dumpToFile(report);
+            deinit();
+            deinitMemory();
+            return false;
+        }
+
+        res = deinit();
+        if (!res)
+        {
+            dumpToFile(report);
+            deinitMemory();
+            return false;
+        }
+
+        deinitMemory();
+
+        report.statsText["rcode"] = "Passed";
+        return true;
+    }
+
+protected:
+
+    cudaDeviceProp devProp;
+    std::auto_ptr<INCVMemAllocator> allocatorGPU;
+    std::auto_ptr<INCVMemAllocator> allocatorCPU;
+
+private:
+
+    std::string testName;
+
+    bool initMemory(NCVTestReport &report)
+    {
+        this->allocatorGPU.reset(new NCVMemStackAllocator(devProp.textureAlignment));
+        this->allocatorCPU.reset(new NCVMemStackAllocator(devProp.textureAlignment));
+
+        if (!this->allocatorGPU.get()->isInitialized() ||
+            !this->allocatorCPU.get()->isInitialized())
+        {
+            report.statsText["rcode"] = "Memory FAILED";
+            return false;
+        }
+
+        if (!this->process())
+        {
+            report.statsText["rcode"] = "Memory FAILED";
+            return false;
+        }
+
+        Ncv32u maxGPUsize = (Ncv32u)this->allocatorGPU.get()->maxSize();
+        Ncv32u maxCPUsize = (Ncv32u)this->allocatorCPU.get()->maxSize();
+
+        report.statsNums["MemGPU"] = maxGPUsize;
+        report.statsNums["MemCPU"] = maxCPUsize;
+
+        this->allocatorGPU.reset(new NCVMemStackAllocator(NCVMemoryTypeDevice, maxGPUsize, devProp.textureAlignment));
+
+        this->allocatorCPU.reset(new NCVMemStackAllocator(NCVMemoryTypeHostPinned, maxCPUsize, devProp.textureAlignment));
+
+        if (!this->allocatorGPU.get()->isInitialized() ||
+            !this->allocatorCPU.get()->isInitialized())
+        {
+            report.statsText["rcode"] = "Memory FAILED";
+            return false;
+        }
+
+        return true;
+    }
+
+    void deinitMemory()
+    {
+        this->allocatorGPU.reset();
+        this->allocatorCPU.reset();
+    }
+
+    void dumpToFile(NCVTestReport &report)
+    {
+        bool bReasonMem = (0 == report.statsText["rcode"].compare("Memory FAILED"));
+        std::string fname = "TestDump_";
+        fname += (bReasonMem ? "m_" : "") + this->testName + ".log";
+        std::ofstream stream(fname.c_str(), std::ios::trunc | std::ios::out);
+        if (!stream.is_open()) return;
+
+        stream << "NCV Test Failure Log: " << this->testName << std::endl;
+        stream << "====================================================" << std::endl << std::endl;
+        stream << "Test initialization report: " << std::endl;
+        for (std::map<std::string,std::string>::iterator it=report.statsText.begin();
+             it != report.statsText.end(); it++)
+        {
+            stream << it->first << "=" << it->second << std::endl;
+        }
+        for (std::map<std::string,Ncv32u>::iterator it=report.statsNums.begin();
+            it != report.statsNums.end(); it++)
+        {
+            stream << it->first << "=" << it->second << std::endl;
+        }
+        stream << std::endl;
+
+        stream << "Test initialization parameters: " << std::endl;
+        bool bSerializeRes = false;
+        try
+        {
+            bSerializeRes = this->toString(stream);
+        }
+        catch (...)
+        {
+        }
+
+        if (!bSerializeRes)
+        {
+            stream << "Couldn't retrieve object dump" << std::endl;
+        }
+
+        stream.flush();
+    }
+};
+
+#endif // _ncvtest_hpp_

modules/gpu/test/nvidia/NCVTestSourceProvider.hpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _ncvtestsourceprovider_hpp_
+#define _ncvtestsourceprovider_hpp_
+
+#include <memory>
+
+#include "NCV.hpp"
+#include <opencv2/highgui/highgui.hpp>
+
+
+template <class T>
+class NCVTestSourceProvider
+{
+public:
+
+    NCVTestSourceProvider(Ncv32u seed, T rangeLow, T rangeHigh, Ncv32u maxWidth, Ncv32u maxHeight)
+        :
+        bInit(false)
+    {
+        ncvAssertPrintReturn(rangeLow < rangeHigh, "NCVTestSourceProvider ctor:: Invalid range", );
+
+        int devId;
+        cudaDeviceProp devProp;
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDevice(&devId), "Error returned from cudaGetDevice", );
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDeviceProperties(&devProp, devId), "Error returned from cudaGetDeviceProperties", );
+
+        //Ncv32u maxWpitch = alignUp(maxWidth * sizeof(T), devProp.textureAlignment);
+
+        allocatorCPU.reset(new NCVMemNativeAllocator(NCVMemoryTypeHostPinned, devProp.textureAlignment));
+        data.reset(new NCVMatrixAlloc<T>(*this->allocatorCPU.get(), maxWidth, maxHeight));
+        ncvAssertPrintReturn(data.get()->isMemAllocated(), "NCVTestSourceProvider ctor:: Matrix not allocated", );
+
+        this->dataWidth = maxWidth;
+        this->dataHeight = maxHeight;
+
+        srand(seed);
+
+        for (Ncv32u i=0; i<maxHeight; i++)
+        {
+            for (Ncv32u j=0; j<data.get()->stride(); j++)
+            {
+                data.get()->ptr()[i * data.get()->stride() + j] =
+                    (T)(((1.0 * rand()) / RAND_MAX) * (rangeHigh - rangeLow) + rangeLow);
+            }
+        }
+
+        this->bInit = true;
+    }
+
+    NCVTestSourceProvider(std::string pgmFilename)
+        :
+        bInit(false)
+    {
+        ncvAssertPrintReturn(sizeof(T) == 1, "NCVTestSourceProvider ctor:: PGM constructor complies only with 8bit types", );
+
+		cv::Mat image = cv::imread(pgmFilename);		
+		ncvAssertPrintReturn(!image.empty(), "NCVTestSourceProvider ctor:: PGM file error", );
+
+        int devId;
+        cudaDeviceProp devProp;
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDevice(&devId), "Error returned from cudaGetDevice", );
+        ncvAssertPrintReturn(cudaSuccess == cudaGetDeviceProperties(&devProp, devId), "Error returned from cudaGetDeviceProperties", );
+
+        allocatorCPU.reset(new NCVMemNativeAllocator(NCVMemoryTypeHostPinned, devProp.textureAlignment));
+        data.reset(new NCVMatrixAlloc<T>(*this->allocatorCPU.get(), image.cols, image.rows));
+        ncvAssertPrintReturn(data.get()->isMemAllocated(), "NCVTestSourceProvider ctor:: Matrix not allocated", );
+
+        this->dataWidth = image.cols;
+        this->dataHeight = image.rows;
+
+		cv::Mat hdr(image.size(), CV_8UC1, data.get()->ptr(), data.get()->pitch());
+		image.copyTo(hdr);
+        
+        this->bInit = true;
+    }
+
+    NcvBool fill(NCVMatrix<T> &dst)
+    {
+        ncvAssertReturn(this->isInit() &&
+                        dst.memType() == allocatorCPU.get()->memType(), false);
+
+        if (dst.width() == 0 || dst.height() == 0)
+        {
+            return true;
+        }
+
+        for (Ncv32u i=0; i<dst.height(); i++)
+        {
+            Ncv32u srcLine = i % this->dataHeight;
+
+            Ncv32u srcFullChunks = dst.width() / this->dataWidth;
+            for (Ncv32u j=0; j<srcFullChunks; j++)
+            {
+                memcpy(dst.ptr() + i * dst.stride() + j * this->dataWidth,
+                    this->data.get()->ptr() + this->data.get()->stride() * srcLine,
+                    this->dataWidth * sizeof(T));
+            }
+
+            Ncv32u srcLastChunk = dst.width() % this->dataWidth;
+            memcpy(dst.ptr() + i * dst.stride() + srcFullChunks * this->dataWidth,
+                this->data.get()->ptr() + this->data.get()->stride() * srcLine,
+                srcLastChunk * sizeof(T));
+        }
+
+        return true;
+    }
+
+    NcvBool fill(NCVVector<T> &dst)
+    {
+        ncvAssertReturn(this->isInit() &&
+                        dst.memType() == allocatorCPU.get()->memType(), false);
+
+        if (dst.length() == 0)
+        {
+            return true;
+        }
+
+        Ncv32u srcLen = this->dataWidth * this->dataHeight;
+
+        Ncv32u srcFullChunks = (Ncv32u)dst.length() / srcLen;
+        for (Ncv32u j=0; j<srcFullChunks; j++)
+        {
+            memcpy(dst.ptr() + j * srcLen, this->data.get()->ptr(), srcLen * sizeof(T));
+        }
+
+        Ncv32u srcLastChunk = dst.length() % srcLen;
+        memcpy(dst.ptr() + srcFullChunks * srcLen, this->data.get()->ptr(), srcLastChunk * sizeof(T));
+
+        return true;
+    }
+
+    ~NCVTestSourceProvider()
+    {
+        data.reset();
+        allocatorCPU.reset();
+    }
+
+private:
+
+    NcvBool isInit(void)
+    {
+        return this->bInit;
+    }
+
+    NcvBool bInit;
+    std::auto_ptr< INCVMemAllocator > allocatorCPU;
+    std::auto_ptr< NCVMatrixAlloc<T> > data;
+    Ncv32u dataWidth;
+    Ncv32u dataHeight;
+};
+
+#endif // _ncvtestsourceprovider_hpp_

modules/gpu/test/nvidia/TestCompact.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestCompact.h"
+
+
+TestCompact::TestCompact(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                                             Ncv32u length, Ncv32u badElem, Ncv32u badElemPercentage)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    length(length),
+    badElem(badElem),
+    badElemPercentage(badElemPercentage > 100 ? 100 : badElemPercentage)
+{
+}
+
+
+bool TestCompact::toString(std::ofstream &strOut)
+{
+    strOut << "length=" << length << std::endl;
+    strOut << "badElem=" << badElem << std::endl;
+    strOut << "badElemPercentage=" << badElemPercentage << std::endl;
+    return true;
+}
+
+
+bool TestCompact::init()
+{
+    return true;
+}
+
+
+bool TestCompact::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    NCVVectorAlloc<Ncv32u> h_vecSrc(*this->allocatorCPU.get(), this->length);
+    ncvAssertReturn(h_vecSrc.isMemAllocated(), false);
+    NCVVectorAlloc<Ncv32u> d_vecSrc(*this->allocatorGPU.get(), this->length);
+    ncvAssertReturn(d_vecSrc.isMemAllocated(), false);
+
+    NCVVectorAlloc<Ncv32u> h_vecDst(*this->allocatorCPU.get(), this->length);
+    ncvAssertReturn(h_vecDst.isMemAllocated(), false);
+    NCVVectorAlloc<Ncv32u> d_vecDst(*this->allocatorGPU.get(), this->length);
+    ncvAssertReturn(d_vecDst.isMemAllocated(), false);
+    NCVVectorAlloc<Ncv32u> h_vecDst_d(*this->allocatorCPU.get(), this->length);
+    ncvAssertReturn(h_vecDst_d.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_vecSrc), false);
+    for (Ncv32u i=0; i<this->length; i++)
+    {
+        Ncv32u tmp = (h_vecSrc.ptr()[i]) & 0xFF;
+        tmp = tmp * 99 / 255;
+        if (tmp < this->badElemPercentage)
+        {
+            h_vecSrc.ptr()[i] = this->badElem;
+        }
+    }
+    NCV_SKIP_COND_END
+
+    NCVVectorAlloc<Ncv32u> h_dstLen(*this->allocatorCPU.get(), 1);
+    ncvAssertReturn(h_dstLen.isMemAllocated(), false);
+    Ncv32u bufSize;
+    ncvStat = nppsStCompactGetSize_32u(this->length, &bufSize, this->devProp);
+    ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    NCVVectorAlloc<Ncv8u> d_tmpBuf(*this->allocatorGPU.get(), bufSize);
+    ncvAssertReturn(d_tmpBuf.isMemAllocated(), false);
+
+    Ncv32u h_outElemNum_h = 0;
+
+    NCV_SKIP_COND_BEGIN
+    ncvStat = h_vecSrc.copySolid(d_vecSrc, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    ncvStat = nppsStCompact_32u(d_vecSrc.ptr(), this->length,
+                                d_vecDst.ptr(), h_dstLen.ptr(), this->badElem,
+                                d_tmpBuf.ptr(), bufSize, this->devProp);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    ncvStat = d_vecDst.copySolid(h_vecDst_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppsStCompact_32u_host(h_vecSrc.ptr(), this->length, h_vecDst.ptr(), &h_outElemNum_h, this->badElem);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    if (h_dstLen.ptr()[0] != h_outElemNum_h)
+    {
+        bLoopVirgin = false;
+    }
+    else
+    {
+        for (Ncv32u i=0; bLoopVirgin && i < h_outElemNum_h; i++)
+        {
+            if (h_vecDst.ptr()[i] != h_vecDst_d.ptr()[i])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestCompact::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestCompact.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testhypothesescompact_h_
+#define _testhypothesescompact_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestCompact : public NCVTestProvider
+{
+public:
+
+    TestCompact(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                          Ncv32u length, Ncv32u badElem, Ncv32u badElemPercentage);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestCompact(const TestCompact&);
+	TestCompact& operator=(const TestCompact&);	
+
+
+    NCVTestSourceProvider<Ncv32u> &src;
+    Ncv32u length;
+    Ncv32u badElem;
+    Ncv32u badElemPercentage;
+};
+
+#endif // _testhypothesescompact_h_

modules/gpu/test/nvidia/TestDrawRects.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestDrawRects.h"
+#include "NCVHaarObjectDetection.hpp"
+
+
+template <class T>
+TestDrawRects<T>::TestDrawRects(std::string testName, NCVTestSourceProvider<T> &src, NCVTestSourceProvider<Ncv32u> &src32u,
+                                Ncv32u width, Ncv32u height, Ncv32u numRects, T color)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    src32u(src32u),
+    width(width),
+    height(height),
+    numRects(numRects),
+    color(color)
+{
+}
+
+
+template <class T>
+bool TestDrawRects<T>::toString(std::ofstream &strOut)
+{
+    strOut << "sizeof(T)=" << sizeof(T) << std::endl;
+    strOut << "width=" << width << std::endl;
+    strOut << "height=" << height << std::endl;
+    strOut << "numRects=" << numRects << std::endl;
+    strOut << "color=" << color << std::endl;
+    return true;
+}
+
+
+template <class T>
+bool TestDrawRects<T>::init()
+{
+    return true;
+}
+
+
+template <class T>
+bool TestDrawRects<T>::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    NCVMatrixAlloc<T> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_img_d(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img_d.isMemAllocated(), false);
+
+    NCVVectorAlloc<NcvRect32u> d_rects(*this->allocatorGPU.get(), this->numRects);
+    ncvAssertReturn(d_rects.isMemAllocated(), false);
+    NCVVectorAlloc<NcvRect32u> h_rects(*this->allocatorCPU.get(), this->numRects);
+    ncvAssertReturn(h_rects.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_img), false);
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    //fill vector of rectangles with random rects covering the input
+    NCVVectorReuse<Ncv32u> h_rects_as32u(h_rects.getSegment());
+    ncvAssertReturn(h_rects_as32u.isMemReused(), false);
+    ncvAssertReturn(this->src32u.fill(h_rects_as32u), false);
+    for (Ncv32u i=0; i<this->numRects; i++)
+    {
+        h_rects.ptr()[i].x = (Ncv32u)(((1.0 * h_rects.ptr()[i].x) / RAND_MAX) * (this->width-2));
+        h_rects.ptr()[i].y = (Ncv32u)(((1.0 * h_rects.ptr()[i].y) / RAND_MAX) * (this->height-2));
+        h_rects.ptr()[i].width = (Ncv32u)(((1.0 * h_rects.ptr()[i].width) / RAND_MAX) * (this->width+10 - h_rects.ptr()[i].x));
+        h_rects.ptr()[i].height = (Ncv32u)(((1.0 * h_rects.ptr()[i].height) / RAND_MAX) * (this->height+10 - h_rects.ptr()[i].y));
+    }
+    ncvStat = h_rects.copySolid(d_rects, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = ncvDrawRects_32u_device((Ncv32u *)d_img.ptr(), d_img.stride(), this->width, this->height,
+                                          (NcvRect32u *)d_rects.ptr(), this->numRects, this->color, 0);
+    }
+    else if (sizeof(T) == sizeof(Ncv8u))
+    {
+        ncvStat = ncvDrawRects_8u_device((Ncv8u *)d_img.ptr(), d_img.stride(), this->width, this->height,
+                                         (NcvRect32u *)d_rects.ptr(), this->numRects, (Ncv8u)this->color, 0);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect drawrects test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    ncvStat = d_img.copySolid(h_img_d, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    NCV_SKIP_COND_BEGIN
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = ncvDrawRects_32u_host((Ncv32u *)h_img.ptr(), h_img.stride(), this->width, this->height,
+                                        (NcvRect32u *)h_rects.ptr(), this->numRects, this->color);
+    }
+    else if (sizeof(T) == sizeof(Ncv8u))
+    {
+        ncvStat = ncvDrawRects_8u_host((Ncv8u *)h_img.ptr(), h_img.stride(), this->width, this->height,
+                                       (NcvRect32u *)h_rects.ptr(), this->numRects, (Ncv8u)this->color);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect drawrects test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    //const Ncv64f relEPS = 0.005;
+    for (Ncv32u i=0; bLoopVirgin && i < h_img.height(); i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < h_img.width(); j++)
+        {
+            if (h_img.ptr()[h_img.stride()*i+j] != h_img_d.ptr()[h_img_d.stride()*i+j])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+template <class T>
+bool TestDrawRects<T>::deinit()
+{
+    return true;
+}
+
+
+template class TestDrawRects<Ncv8u>;
+template class TestDrawRects<Ncv32u>;

modules/gpu/test/nvidia/TestDrawRects.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testdrawrects_h_
+#define _testdrawrects_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+template <class T>
+class TestDrawRects : public NCVTestProvider
+{
+public:
+
+    TestDrawRects(std::string testName, NCVTestSourceProvider<T> &src, NCVTestSourceProvider<Ncv32u> &src32u,
+                  Ncv32u width, Ncv32u height, Ncv32u numRects, T color);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+
+	TestDrawRects(const TestDrawRects&);
+	TestDrawRects& operator=(const TestDrawRects&);	
+
+    NCVTestSourceProvider<T> &src;
+    NCVTestSourceProvider<Ncv32u> &src32u;
+    Ncv32u width;
+    Ncv32u height;
+    Ncv32u numRects;
+    T color;
+};
+
+#endif // _testdrawrects_h_

modules/gpu/test/nvidia/TestHaarCascadeApplication.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include <float.h>
+
+#if defined(__GNUC__)
+    #include <fpu_control.h>
+#endif
+
+#include "TestHaarCascadeApplication.h"
+#include "NCVHaarObjectDetection.hpp"
+
+
+TestHaarCascadeApplication::TestHaarCascadeApplication(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                                                       std::string cascadeName, Ncv32u width, Ncv32u height)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    cascadeName(cascadeName),
+    width(width),
+    height(height)
+{
+}
+
+
+bool TestHaarCascadeApplication::toString(std::ofstream &strOut)
+{
+    strOut << "cascadeName=" << cascadeName << std::endl;
+    strOut << "width=" << width << std::endl;
+    strOut << "height=" << height << std::endl;
+    return true;
+}
+
+
+bool TestHaarCascadeApplication::init()
+{
+    return true;
+}
+
+
+bool TestHaarCascadeApplication::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32u numStages, numNodes, numFeatures;
+
+    ncvStat = ncvHaarGetClassifierSize(this->cascadeName, numStages, numNodes, numFeatures);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    NCVVectorAlloc<HaarStage64> h_HaarStages(*this->allocatorCPU.get(), numStages);
+    ncvAssertReturn(h_HaarStages.isMemAllocated(), false);
+    NCVVectorAlloc<HaarClassifierNode128> h_HaarNodes(*this->allocatorCPU.get(), numNodes);
+    ncvAssertReturn(h_HaarNodes.isMemAllocated(), false);
+    NCVVectorAlloc<HaarFeature64> h_HaarFeatures(*this->allocatorCPU.get(), numFeatures);
+    ncvAssertReturn(h_HaarFeatures.isMemAllocated(), false);
+
+    NCVVectorAlloc<HaarStage64> d_HaarStages(*this->allocatorGPU.get(), numStages);
+    ncvAssertReturn(d_HaarStages.isMemAllocated(), false);
+    NCVVectorAlloc<HaarClassifierNode128> d_HaarNodes(*this->allocatorGPU.get(), numNodes);
+    ncvAssertReturn(d_HaarNodes.isMemAllocated(), false);
+    NCVVectorAlloc<HaarFeature64> d_HaarFeatures(*this->allocatorGPU.get(), numFeatures);
+    ncvAssertReturn(d_HaarFeatures.isMemAllocated(), false);
+
+    HaarClassifierCascadeDescriptor haar;
+    haar.ClassifierSize.width = haar.ClassifierSize.height = 1;
+    haar.bNeedsTiltedII = false;
+    haar.NumClassifierRootNodes = numNodes;
+    haar.NumClassifierTotalNodes = numNodes;
+    haar.NumFeatures = numFeatures;
+    haar.NumStages = numStages;
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+
+    ncvStat = ncvHaarLoadFromFile_host(this->cascadeName, haar, h_HaarStages, h_HaarNodes, h_HaarFeatures);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    ncvAssertReturn(NCV_SUCCESS == h_HaarStages.copySolid(d_HaarStages, 0), false);
+    ncvAssertReturn(NCV_SUCCESS == h_HaarNodes.copySolid(d_HaarNodes, 0), false);
+    ncvAssertReturn(NCV_SUCCESS == h_HaarFeatures.copySolid(d_HaarFeatures, 0), false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    NCV_SKIP_COND_END
+
+    NcvSize32s srcRoi, srcIIRoi, searchRoi;
+    srcRoi.width = this->width;
+    srcRoi.height = this->height;
+    srcIIRoi.width = srcRoi.width + 1;
+    srcIIRoi.height = srcRoi.height + 1;
+    searchRoi.width = srcIIRoi.width - haar.ClassifierSize.width;
+    searchRoi.height = srcIIRoi.height - haar.ClassifierSize.height;
+    if (searchRoi.width <= 0 || searchRoi.height <= 0)
+    {
+        return false;
+    }
+    NcvSize32u searchRoiU(searchRoi.width, searchRoi.height);
+
+    NCVMatrixAlloc<Ncv8u> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv8u> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+
+    Ncv32u integralWidth = this->width + 1;
+    Ncv32u integralHeight = this->height + 1;
+
+    NCVMatrixAlloc<Ncv32u> d_integralImage(*this->allocatorGPU.get(), integralWidth, integralHeight);
+    ncvAssertReturn(d_integralImage.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> d_sqIntegralImage(*this->allocatorGPU.get(), integralWidth, integralHeight);
+    ncvAssertReturn(d_sqIntegralImage.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32u> h_integralImage(*this->allocatorCPU.get(), integralWidth, integralHeight);
+    ncvAssertReturn(h_integralImage.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> h_sqIntegralImage(*this->allocatorCPU.get(), integralWidth, integralHeight);
+    ncvAssertReturn(h_sqIntegralImage.isMemAllocated(), false);
+
+    NCVMatrixAlloc<Ncv32f> d_rectStdDev(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_rectStdDev.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32u> d_pixelMask(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_pixelMask.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32f> h_rectStdDev(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_rectStdDev.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32u> h_pixelMask(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_pixelMask.isMemAllocated(), false);
+
+    NCVVectorAlloc<NcvRect32u> d_hypotheses(*this->allocatorGPU.get(), this->width * this->height);
+    ncvAssertReturn(d_hypotheses.isMemAllocated(), false);
+    NCVVectorAlloc<NcvRect32u> h_hypotheses(*this->allocatorCPU.get(), this->width * this->height);
+    ncvAssertReturn(h_hypotheses.isMemAllocated(), false);
+
+    NCVStatus nppStat;
+    Ncv32u szTmpBufIntegral, szTmpBufSqIntegral;
+    nppStat = nppiStIntegralGetSize_8u32u(NcvSize32u(this->width, this->height), &szTmpBufIntegral, this->devProp);
+    ncvAssertReturn(nppStat == NPPST_SUCCESS, false);
+    nppStat = nppiStSqrIntegralGetSize_8u64u(NcvSize32u(this->width, this->height), &szTmpBufSqIntegral, this->devProp);
+    ncvAssertReturn(nppStat == NPPST_SUCCESS, false);
+    NCVVectorAlloc<Ncv8u> d_tmpIIbuf(*this->allocatorGPU.get(), std::max(szTmpBufIntegral, szTmpBufSqIntegral));
+    ncvAssertReturn(d_tmpIIbuf.isMemAllocated(), false);
+
+    Ncv32u detectionsOnThisScale_d = 0;
+    Ncv32u detectionsOnThisScale_h = 0;
+
+    NCV_SKIP_COND_BEGIN
+
+    ncvAssertReturn(this->src.fill(h_img), false);
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    nppStat = nppiStIntegral_8u32u_C1R(d_img.ptr(), d_img.pitch(),
+                                       d_integralImage.ptr(), d_integralImage.pitch(),
+                                       NcvSize32u(d_img.width(), d_img.height()),
+                                       d_tmpIIbuf.ptr(), szTmpBufIntegral, this->devProp);
+    ncvAssertReturn(nppStat == NPPST_SUCCESS, false);
+
+    nppStat = nppiStSqrIntegral_8u64u_C1R(d_img.ptr(), d_img.pitch(),
+                                          d_sqIntegralImage.ptr(), d_sqIntegralImage.pitch(),
+                                          NcvSize32u(d_img.width(), d_img.height()),
+                                          d_tmpIIbuf.ptr(), szTmpBufSqIntegral, this->devProp);
+    ncvAssertReturn(nppStat == NPPST_SUCCESS, false);
+
+    const NcvRect32u rect(
+        HAAR_STDDEV_BORDER,
+        HAAR_STDDEV_BORDER,
+        haar.ClassifierSize.width - 2*HAAR_STDDEV_BORDER,
+        haar.ClassifierSize.height - 2*HAAR_STDDEV_BORDER);
+    nppStat = nppiStRectStdDev_32f_C1R(
+        d_integralImage.ptr(), d_integralImage.pitch(),
+        d_sqIntegralImage.ptr(), d_sqIntegralImage.pitch(),
+        d_rectStdDev.ptr(), d_rectStdDev.pitch(),
+        NcvSize32u(searchRoi.width, searchRoi.height), rect,
+        1.0f, true);
+    ncvAssertReturn(nppStat == NPPST_SUCCESS, false);
+
+    ncvStat = d_integralImage.copySolid(h_integralImage, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvStat = d_rectStdDev.copySolid(h_rectStdDev, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    for (Ncv32u i=0; i<searchRoiU.height; i++)
+    {
+        for (Ncv32u j=0; j<h_pixelMask.stride(); j++)
+        {
+            if (j<searchRoiU.width)
+            {
+                h_pixelMask.ptr()[i*h_pixelMask.stride()+j] = (i << 16) | j;
+            }
+            else
+            {
+                h_pixelMask.ptr()[i*h_pixelMask.stride()+j] = OBJDET_MASK_ELEMENT_INVALID_32U;
+            }
+        }
+    }
+    ncvAssertReturn(cudaSuccess == cudaStreamSynchronize(0), false);
+
+#if defined(__GNUC__)
+    //http://www.christian-seiler.de/projekte/fpmath/
+
+    fpu_control_t fpu_oldcw, fpu_cw;
+    _FPU_GETCW(fpu_oldcw); // store old cw
+     fpu_cw = (fpu_oldcw & ~_FPU_EXTENDED & ~_FPU_DOUBLE & ~_FPU_SINGLE) | _FPU_SINGLE;
+    _FPU_SETCW(fpu_cw);
+
+    // calculations here
+    ncvStat = ncvApplyHaarClassifierCascade_host(
+        h_integralImage, h_rectStdDev, h_pixelMask,
+        detectionsOnThisScale_h,
+        haar, h_HaarStages, h_HaarNodes, h_HaarFeatures, false,
+        searchRoiU, 1, 1.0f);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    _FPU_SETCW(fpu_oldcw); // restore old cw
+#else
+    Ncv32u fpu_oldcw, fpu_cw;
+    _controlfp_s(&fpu_cw, 0, 0);
+    fpu_oldcw = fpu_cw;
+    _controlfp_s(&fpu_cw, _PC_24, _MCW_PC);
+    ncvStat = ncvApplyHaarClassifierCascade_host(
+        h_integralImage, h_rectStdDev, h_pixelMask,
+        detectionsOnThisScale_h,
+        haar, h_HaarStages, h_HaarNodes, h_HaarFeatures, false,
+        searchRoiU, 1, 1.0f);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    _controlfp_s(&fpu_cw, fpu_oldcw, _MCW_PC);
+#endif
+    NCV_SKIP_COND_END
+
+    int devId;
+    ncvAssertCUDAReturn(cudaGetDevice(&devId), false);
+    cudaDeviceProp devProp;
+    ncvAssertCUDAReturn(cudaGetDeviceProperties(&devProp, devId), false);
+
+    ncvStat = ncvApplyHaarClassifierCascade_device(
+        d_integralImage, d_rectStdDev, d_pixelMask,
+        detectionsOnThisScale_d,
+        haar, h_HaarStages, d_HaarStages, d_HaarNodes, d_HaarFeatures, false,
+        searchRoiU, 1, 1.0f,
+        *this->allocatorGPU.get(), *this->allocatorCPU.get(),
+        devProp, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    NCVMatrixAlloc<Ncv32u> h_pixelMask_d(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_pixelMask_d.isMemAllocated(), false);
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+
+    ncvStat = d_pixelMask.copySolid(h_pixelMask_d, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    if (detectionsOnThisScale_d != detectionsOnThisScale_h)
+    {
+        bLoopVirgin = false;
+    }
+    else
+    {
+        std::sort(h_pixelMask_d.ptr(), h_pixelMask_d.ptr() + detectionsOnThisScale_d);
+        for (Ncv32u i=0; i<detectionsOnThisScale_d && bLoopVirgin; i++)
+        {
+            if (h_pixelMask.ptr()[i] != h_pixelMask_d.ptr()[i])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestHaarCascadeApplication::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestHaarCascadeApplication.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testhaarcascadeapplication_h_
+#define _testhaarcascadeapplication_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestHaarCascadeApplication : public NCVTestProvider
+{
+public:
+
+    TestHaarCascadeApplication(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                               std::string cascadeName, Ncv32u width, Ncv32u height);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestHaarCascadeApplication(const TestHaarCascadeApplication&);
+	TestHaarCascadeApplication& operator=(const TestHaarCascadeApplication&);	
+
+
+    NCVTestSourceProvider<Ncv8u> &src;
+    std::string cascadeName;
+    Ncv32u width;
+    Ncv32u height;
+};
+
+#endif // _testhaarcascadeapplication_h_

modules/gpu/test/nvidia/TestHaarCascadeLoader.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestHaarCascadeLoader.h"
+#include "NCVHaarObjectDetection.hpp"
+
+
+TestHaarCascadeLoader::TestHaarCascadeLoader(std::string testName, std::string cascadeName)
+    :
+    NCVTestProvider(testName),
+    cascadeName(cascadeName)
+{
+}
+
+
+bool TestHaarCascadeLoader::toString(std::ofstream &strOut)
+{
+    strOut << "cascadeName=" << cascadeName << std::endl;
+    return true;
+}
+
+
+bool TestHaarCascadeLoader::init()
+{
+    return true;
+}
+
+
+bool TestHaarCascadeLoader::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32u numStages, numNodes, numFeatures;
+    Ncv32u numStages_2 = 0, numNodes_2 = 0, numFeatures_2 = 0;
+
+    ncvStat = ncvHaarGetClassifierSize(this->cascadeName, numStages, numNodes, numFeatures);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    NCVVectorAlloc<HaarStage64> h_HaarStages(*this->allocatorCPU.get(), numStages);
+    ncvAssertReturn(h_HaarStages.isMemAllocated(), false);
+    NCVVectorAlloc<HaarClassifierNode128> h_HaarNodes(*this->allocatorCPU.get(), numNodes);
+    ncvAssertReturn(h_HaarNodes.isMemAllocated(), false);
+    NCVVectorAlloc<HaarFeature64> h_HaarFeatures(*this->allocatorCPU.get(), numFeatures);
+    ncvAssertReturn(h_HaarFeatures.isMemAllocated(), false);
+
+    NCVVectorAlloc<HaarStage64> h_HaarStages_2(*this->allocatorCPU.get(), numStages);
+    ncvAssertReturn(h_HaarStages_2.isMemAllocated(), false);
+    NCVVectorAlloc<HaarClassifierNode128> h_HaarNodes_2(*this->allocatorCPU.get(), numNodes);
+    ncvAssertReturn(h_HaarNodes_2.isMemAllocated(), false);
+    NCVVectorAlloc<HaarFeature64> h_HaarFeatures_2(*this->allocatorCPU.get(), numFeatures);
+    ncvAssertReturn(h_HaarFeatures_2.isMemAllocated(), false);
+
+    HaarClassifierCascadeDescriptor haar;
+    HaarClassifierCascadeDescriptor haar_2;
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+
+    const std::string testNvbinName = "test.nvbin";
+    ncvStat = ncvHaarLoadFromFile_host(this->cascadeName, haar, h_HaarStages, h_HaarNodes, h_HaarFeatures);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    ncvStat = ncvHaarStoreNVBIN_host(testNvbinName, haar, h_HaarStages, h_HaarNodes, h_HaarFeatures);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    ncvStat = ncvHaarGetClassifierSize(testNvbinName, numStages_2, numNodes_2, numFeatures_2);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    ncvStat = ncvHaarLoadFromFile_host(testNvbinName, haar_2, h_HaarStages_2, h_HaarNodes_2, h_HaarFeatures_2);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+
+    if (
+    numStages_2 != numStages                                       ||
+    numNodes_2 != numNodes                                         ||
+    numFeatures_2 != numFeatures                                   ||
+    haar.NumStages               != haar_2.NumStages               ||
+    haar.NumClassifierRootNodes  != haar_2.NumClassifierRootNodes  ||
+    haar.NumClassifierTotalNodes != haar_2.NumClassifierTotalNodes ||
+    haar.NumFeatures             != haar_2.NumFeatures             ||
+    haar.ClassifierSize.width    != haar_2.ClassifierSize.width    ||
+    haar.ClassifierSize.height   != haar_2.ClassifierSize.height   ||
+    haar.bNeedsTiltedII          != haar_2.bNeedsTiltedII          ||
+    haar.bHasStumpsOnly          != haar_2.bHasStumpsOnly          )
+    {
+        bLoopVirgin = false;
+    }
+    if (memcmp(h_HaarStages.ptr(), h_HaarStages_2.ptr(), haar.NumStages * sizeof(HaarStage64)) ||
+        memcmp(h_HaarNodes.ptr(), h_HaarNodes_2.ptr(), haar.NumClassifierTotalNodes * sizeof(HaarClassifierNode128)) ||
+        memcmp(h_HaarFeatures.ptr(), h_HaarFeatures_2.ptr(), haar.NumFeatures * sizeof(HaarFeature64)) )
+    {
+        bLoopVirgin = false;
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestHaarCascadeLoader::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestHaarCascadeLoader.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testhaarcascadeloader_h_
+#define _testhaarcascadeloader_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestHaarCascadeLoader : public NCVTestProvider
+{
+public:
+
+    TestHaarCascadeLoader(std::string testName, std::string cascadeName);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+
+    std::string cascadeName;
+};
+
+#endif // _testhaarcascadeloader_h_

modules/gpu/test/nvidia/TestHypothesesFilter.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestHypothesesFilter.h"
+#include "NCVHaarObjectDetection.hpp"
+
+
+TestHypothesesFilter::TestHypothesesFilter(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                                           Ncv32u numDstRects, Ncv32u minNeighbors, Ncv32f eps)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    numDstRects(numDstRects),
+    minNeighbors(minNeighbors),
+    eps(eps)
+{
+}
+
+
+bool TestHypothesesFilter::toString(std::ofstream &strOut)
+{
+    strOut << "numDstRects=" << numDstRects << std::endl;
+    strOut << "minNeighbors=" << minNeighbors << std::endl;
+    strOut << "eps=" << eps << std::endl;
+    return true;
+}
+
+
+bool TestHypothesesFilter::init()
+{
+    this->canvasWidth = 4096;
+    this->canvasHeight = 4096;
+    return true;
+}
+
+
+bool compareRects(const NcvRect32u &r1, const NcvRect32u &r2, Ncv32f eps)
+{
+    double delta = eps*(std::min(r1.width, r2.width) + std::min(r1.height, r2.height))*0.5;
+    return std::abs((Ncv32s)r1.x - (Ncv32s)r2.x) <= delta &&
+        std::abs((Ncv32s)r1.y - (Ncv32s)r2.y) <= delta &&
+        std::abs((Ncv32s)r1.x + (Ncv32s)r1.width - (Ncv32s)r2.x - (Ncv32s)r2.width) <= delta &&
+        std::abs((Ncv32s)r1.y + (Ncv32s)r1.height - (Ncv32s)r2.y - (Ncv32s)r2.height) <= delta;
+}
+
+
+inline bool operator < (const NcvRect32u &a, const NcvRect32u &b)
+{
+    return a.x < b.x;
+}
+
+
+bool TestHypothesesFilter::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    NCVVectorAlloc<Ncv32u> h_random32u(*this->allocatorCPU.get(), this->numDstRects * sizeof(NcvRect32u) / sizeof(Ncv32u));
+    ncvAssertReturn(h_random32u.isMemAllocated(), false);
+
+    Ncv32u srcSlotSize = 2 * this->minNeighbors + 1;
+
+    NCVVectorAlloc<NcvRect32u> h_vecSrc(*this->allocatorCPU.get(), this->numDstRects*srcSlotSize);
+    ncvAssertReturn(h_vecSrc.isMemAllocated(), false);
+    NCVVectorAlloc<NcvRect32u> h_vecDst_groundTruth(*this->allocatorCPU.get(), this->numDstRects);
+    ncvAssertReturn(h_vecDst_groundTruth.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorCPU.get()->isCounting());
+
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_random32u), false);
+    Ncv32u randCnt = 0;
+    Ncv64f randVal;
+
+    for (Ncv32u i=0; i<this->numDstRects; i++)
+    {
+        h_vecDst_groundTruth.ptr()[i].x = i * this->canvasWidth / this->numDstRects + this->canvasWidth / (this->numDstRects * 4);
+        h_vecDst_groundTruth.ptr()[i].y = i * this->canvasHeight / this->numDstRects + this->canvasHeight / (this->numDstRects * 4);
+        h_vecDst_groundTruth.ptr()[i].width = this->canvasWidth / (this->numDstRects * 2);
+        h_vecDst_groundTruth.ptr()[i].height = this->canvasHeight / (this->numDstRects * 2);
+
+        Ncv32u numNeighbors = this->minNeighbors + 1 + (Ncv32u)(((1.0 * h_random32u.ptr()[i]) * (this->minNeighbors + 1)) / 0xFFFFFFFF);
+        numNeighbors = (numNeighbors > srcSlotSize) ? srcSlotSize : numNeighbors;
+
+        //fill in strong hypotheses                           (2 * ((1.0 * randVal) / 0xFFFFFFFF) - 1)
+        for (Ncv32u j=0; j<numNeighbors; j++)
+        {
+            randVal = (1.0 * h_random32u.ptr()[randCnt++]) / 0xFFFFFFFF; randCnt = randCnt % h_random32u.length();
+            h_vecSrc.ptr()[srcSlotSize * i + j].x = 
+                h_vecDst_groundTruth.ptr()[i].x +
+                (Ncv32s)(h_vecDst_groundTruth.ptr()[i].width * this->eps * (randVal - 0.5));
+            randVal = (1.0 * h_random32u.ptr()[randCnt++]) / 0xFFFFFFFF; randCnt = randCnt % h_random32u.length();
+            h_vecSrc.ptr()[srcSlotSize * i + j].y = 
+                h_vecDst_groundTruth.ptr()[i].y +
+                (Ncv32s)(h_vecDst_groundTruth.ptr()[i].height * this->eps * (randVal - 0.5));
+            h_vecSrc.ptr()[srcSlotSize * i + j].width = h_vecDst_groundTruth.ptr()[i].width;
+            h_vecSrc.ptr()[srcSlotSize * i + j].height = h_vecDst_groundTruth.ptr()[i].height;
+        }
+
+        //generate weak hypotheses (to be removed in processing)
+        for (Ncv32u j=numNeighbors; j<srcSlotSize; j++)
+        {
+            randVal = (1.0 * h_random32u.ptr()[randCnt++]) / 0xFFFFFFFF; randCnt = randCnt % h_random32u.length();
+            h_vecSrc.ptr()[srcSlotSize * i + j].x = 
+                this->canvasWidth + h_vecDst_groundTruth.ptr()[i].x +
+                (Ncv32s)(h_vecDst_groundTruth.ptr()[i].width * this->eps * (randVal - 0.5));
+            randVal = (1.0 * h_random32u.ptr()[randCnt++]) / 0xFFFFFFFF; randCnt = randCnt % h_random32u.length();
+            h_vecSrc.ptr()[srcSlotSize * i + j].y = 
+                this->canvasHeight + h_vecDst_groundTruth.ptr()[i].y +
+                (Ncv32s)(h_vecDst_groundTruth.ptr()[i].height * this->eps * (randVal - 0.5));
+            h_vecSrc.ptr()[srcSlotSize * i + j].width = h_vecDst_groundTruth.ptr()[i].width;
+            h_vecSrc.ptr()[srcSlotSize * i + j].height = h_vecDst_groundTruth.ptr()[i].height;
+        }
+    }
+
+    //shuffle
+    for (Ncv32u i=0; i<this->numDstRects*srcSlotSize-1; i++)
+    {
+        Ncv32u randVal = h_random32u.ptr()[randCnt++]; randCnt = randCnt % h_random32u.length();
+        Ncv32u secondSwap = randVal % (this->numDstRects*srcSlotSize-1 - i);
+        NcvRect32u tmp = h_vecSrc.ptr()[i + secondSwap];
+        h_vecSrc.ptr()[i + secondSwap] = h_vecSrc.ptr()[i];
+        h_vecSrc.ptr()[i] = tmp;
+    }
+    NCV_SKIP_COND_END
+
+    Ncv32u numHypothesesSrc = h_vecSrc.length();
+    NCV_SKIP_COND_BEGIN
+    ncvStat = ncvFilterHypotheses_host(h_vecSrc, numHypothesesSrc, this->minNeighbors, this->eps, NULL);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //verification
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    if (numHypothesesSrc != this->numDstRects)
+    {
+        bLoopVirgin = false;
+    }
+    else
+    {
+        std::vector<NcvRect32u> tmpRects(numHypothesesSrc);
+        memcpy(&tmpRects[0], h_vecSrc.ptr(), numHypothesesSrc * sizeof(NcvRect32u));
+        std::sort(tmpRects.begin(), tmpRects.end());
+        for (Ncv32u i=0; i<numHypothesesSrc && bLoopVirgin; i++)
+        {
+            if (!compareRects(tmpRects[i], h_vecDst_groundTruth.ptr()[i], this->eps))
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestHypothesesFilter::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestHypothesesFilter.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testhypothesesfilter_h_
+#define _testhypothesesfilter_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestHypothesesFilter : public NCVTestProvider
+{
+public:
+
+    TestHypothesesFilter(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                         Ncv32u numDstRects, Ncv32u minNeighbors, Ncv32f eps);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+
+	TestHypothesesFilter(const TestHypothesesFilter&);
+	TestHypothesesFilter& operator=(const TestHypothesesFilter&);	
+
+    NCVTestSourceProvider<Ncv32u> &src;
+    Ncv32u numDstRects;
+    Ncv32u minNeighbors;
+    Ncv32f eps;
+
+    Ncv32u canvasWidth;
+    Ncv32u canvasHeight;
+};
+
+#endif // _testhypothesesfilter_h_

modules/gpu/test/nvidia/TestHypothesesGrow.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestHypothesesGrow.h"
+#include "NCVHaarObjectDetection.hpp"
+
+
+TestHypothesesGrow::TestHypothesesGrow(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                                       Ncv32u rectWidth, Ncv32u rectHeight, Ncv32f rectScale, 
+                                       Ncv32u maxLenSrc, Ncv32u lenSrc, Ncv32u maxLenDst, Ncv32u lenDst)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    rectWidth(rectWidth),
+    rectHeight(rectHeight),
+    rectScale(rectScale),
+    maxLenSrc(maxLenSrc),
+    lenSrc(lenSrc),
+    maxLenDst(maxLenDst),
+    lenDst(lenDst)
+{
+}
+
+
+bool TestHypothesesGrow::toString(std::ofstream &strOut)
+{
+    strOut << "rectWidth=" << rectWidth << std::endl;
+    strOut << "rectHeight=" << rectHeight << std::endl;
+    strOut << "rectScale=" << rectScale << std::endl;
+    strOut << "maxLenSrc=" << maxLenSrc << std::endl;
+    strOut << "lenSrc=" << lenSrc << std::endl;
+    strOut << "maxLenDst=" << maxLenDst << std::endl;
+    strOut << "lenDst=" << lenDst << std::endl;
+    return true;
+}
+
+
+bool TestHypothesesGrow::init()
+{
+    return true;
+}
+
+
+bool TestHypothesesGrow::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    NCVVectorAlloc<Ncv32u> h_vecSrc(*this->allocatorCPU.get(), this->maxLenSrc);
+    ncvAssertReturn(h_vecSrc.isMemAllocated(), false);
+    NCVVectorAlloc<Ncv32u> d_vecSrc(*this->allocatorGPU.get(), this->maxLenSrc);
+    ncvAssertReturn(d_vecSrc.isMemAllocated(), false);
+
+    NCVVectorAlloc<NcvRect32u> h_vecDst(*this->allocatorCPU.get(), this->maxLenDst);
+    ncvAssertReturn(h_vecDst.isMemAllocated(), false);
+    NCVVectorAlloc<NcvRect32u> d_vecDst(*this->allocatorGPU.get(), this->maxLenDst);
+    ncvAssertReturn(d_vecDst.isMemAllocated(), false);
+    NCVVectorAlloc<NcvRect32u> h_vecDst_d(*this->allocatorCPU.get(), this->maxLenDst);
+    ncvAssertReturn(h_vecDst_d.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_vecSrc), false);
+    memset(h_vecDst.ptr(), 0, h_vecDst.length() * sizeof(NcvRect32u));
+    NCVVectorReuse<Ncv32u> h_vecDst_as32u(h_vecDst.getSegment(), lenDst * sizeof(NcvRect32u) / sizeof(Ncv32u));
+    ncvAssertReturn(h_vecDst_as32u.isMemReused(), false);
+    ncvAssertReturn(this->src.fill(h_vecDst_as32u), false);
+    memcpy(h_vecDst_d.ptr(), h_vecDst.ptr(), h_vecDst.length() * sizeof(NcvRect32u));
+    NCV_SKIP_COND_END
+
+    ncvStat = h_vecSrc.copySolid(d_vecSrc, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvStat = h_vecDst.copySolid(d_vecDst, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    Ncv32u h_outElemNum_d = 0;
+    Ncv32u h_outElemNum_h = 0;
+    NCV_SKIP_COND_BEGIN
+    h_outElemNum_d = this->lenDst;
+    ncvStat = ncvGrowDetectionsVector_device(d_vecSrc, this->lenSrc,
+                                             d_vecDst, h_outElemNum_d, this->maxLenDst,
+                                             this->rectWidth, this->rectHeight, this->rectScale, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvStat = d_vecDst.copySolid(h_vecDst_d, 0);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    ncvAssertCUDAReturn(cudaStreamSynchronize(0), false);
+
+    h_outElemNum_h = this->lenDst;
+    ncvStat = ncvGrowDetectionsVector_host(h_vecSrc, this->lenSrc,
+                                           h_vecDst, h_outElemNum_h, this->maxLenDst,
+                                           this->rectWidth, this->rectHeight, this->rectScale);
+    ncvAssertReturn(ncvStat == NCV_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    if (h_outElemNum_d != h_outElemNum_h)
+    {
+        bLoopVirgin = false;
+    }
+    else
+    {
+        if (memcmp(h_vecDst.ptr(), h_vecDst_d.ptr(), this->maxLenDst * sizeof(NcvRect32u)))
+        {
+            bLoopVirgin = false;
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestHypothesesGrow::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestHypothesesGrow.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testhypothesesgrow_h_
+#define _testhypothesesgrow_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestHypothesesGrow : public NCVTestProvider
+{
+public:
+
+    TestHypothesesGrow(std::string testName, NCVTestSourceProvider<Ncv32u> &src,
+                       Ncv32u rectWidth, Ncv32u rectHeight, Ncv32f rectScale, 
+                       Ncv32u maxLenSrc, Ncv32u lenSrc, Ncv32u maxLenDst, Ncv32u lenDst);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestHypothesesGrow(const TestHypothesesGrow&);
+	TestHypothesesGrow& operator=(const TestHypothesesGrow&);	
+
+
+    NCVTestSourceProvider<Ncv32u> &src;
+    Ncv32u rectWidth;
+    Ncv32u rectHeight;
+    Ncv32f rectScale;
+    Ncv32u maxLenSrc;
+    Ncv32u lenSrc;
+    Ncv32u maxLenDst;
+    Ncv32u lenDst;
+};
+
+#endif // _testhypothesesgrow_h_

modules/gpu/test/nvidia/TestIntegralImage.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include <math.h>
+#include "TestIntegralImage.h"
+
+
+template <class T_in, class T_out>
+TestIntegralImage<T_in, T_out>::TestIntegralImage(std::string testName, NCVTestSourceProvider<T_in> &src,
+                                                  Ncv32u width, Ncv32u height)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    width(width),
+    height(height)
+{
+}
+
+
+template <class T_in, class T_out>
+bool TestIntegralImage<T_in, T_out>::toString(std::ofstream &strOut)
+{
+    strOut << "sizeof(T_in)=" << sizeof(T_in) << std::endl;
+    strOut << "sizeof(T_out)=" << sizeof(T_out) << std::endl;
+    strOut << "width=" << width << std::endl;
+    strOut << "height=" << height << std::endl;
+    return true;
+}
+
+
+template <class T_in, class T_out>
+bool TestIntegralImage<T_in, T_out>::init()
+{
+    return true;
+}
+
+
+template <class T_in, class T_out>
+bool TestIntegralImage<T_in, T_out>::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32u widthII = this->width + 1;
+    Ncv32u heightII = this->height + 1;
+
+    NCVMatrixAlloc<T_in> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T_in> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T_out> d_imgII(*this->allocatorGPU.get(), widthII, heightII);
+    ncvAssertReturn(d_imgII.isMemAllocated(), false);
+    NCVMatrixAlloc<T_out> h_imgII(*this->allocatorCPU.get(), widthII, heightII);
+    ncvAssertReturn(h_imgII.isMemAllocated(), false);
+    NCVMatrixAlloc<T_out> h_imgII_d(*this->allocatorCPU.get(), widthII, heightII);
+    ncvAssertReturn(h_imgII_d.isMemAllocated(), false);
+
+    Ncv32u bufSize;
+    if (sizeof(T_in) == sizeof(Ncv8u))
+    {
+        ncvStat = nppiStIntegralGetSize_8u32u(NcvSize32u(this->width, this->height), &bufSize, this->devProp);
+        ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    }
+    else if (sizeof(T_in) == sizeof(Ncv32f))
+    {
+        ncvStat = nppiStIntegralGetSize_32f32f(NcvSize32u(this->width, this->height), &bufSize, this->devProp);
+        ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect integral image test instance", false);
+    }
+
+    NCVVectorAlloc<Ncv8u> d_tmpBuf(*this->allocatorGPU.get(), bufSize);
+    ncvAssertReturn(d_tmpBuf.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+
+    ncvAssertReturn(this->src.fill(h_img), false);
+
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    if (sizeof(T_in) == sizeof(Ncv8u))
+    {
+        ncvStat = nppiStIntegral_8u32u_C1R((Ncv8u *)d_img.ptr(), d_img.pitch(),
+                                           (Ncv32u *)d_imgII.ptr(), d_imgII.pitch(),
+                                           NcvSize32u(this->width, this->height),
+                                           d_tmpBuf.ptr(), bufSize, this->devProp);
+        ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    }
+    else if (sizeof(T_in) == sizeof(Ncv32f))
+    {
+        ncvStat = nppiStIntegral_32f32f_C1R((Ncv32f *)d_img.ptr(), d_img.pitch(),
+                                            (Ncv32f *)d_imgII.ptr(), d_imgII.pitch(),
+                                            NcvSize32u(this->width, this->height),
+                                            d_tmpBuf.ptr(), bufSize, this->devProp);
+        ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect integral image test instance", false);
+    }
+
+    ncvStat = d_imgII.copySolid(h_imgII_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    if (sizeof(T_in) == sizeof(Ncv8u))
+    {
+        ncvStat = nppiStIntegral_8u32u_C1R_host((Ncv8u *)h_img.ptr(), h_img.pitch(),
+                                                (Ncv32u *)h_imgII.ptr(), h_imgII.pitch(),
+                                                NcvSize32u(this->width, this->height));
+        ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    }
+    else if (sizeof(T_in) == sizeof(Ncv32f))
+    {
+        ncvStat = nppiStIntegral_32f32f_C1R_host((Ncv32f *)h_img.ptr(), h_img.pitch(),
+                                                 (Ncv32f *)h_imgII.ptr(), h_imgII.pitch(),
+                                                 NcvSize32u(this->width, this->height));
+        ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect integral image test instance", false);
+    }
+
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    for (Ncv32u i=0; bLoopVirgin && i < h_img.height() + 1; i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < h_img.width() + 1; j++)
+        {
+            if (sizeof(T_in) == sizeof(Ncv8u))
+            {
+                if (h_imgII.ptr()[h_imgII.stride()*i+j] != h_imgII_d.ptr()[h_imgII_d.stride()*i+j])
+                {
+                    bLoopVirgin = false;
+                }
+            }
+            else if (sizeof(T_in) == sizeof(Ncv32f))
+            {
+                if (fabsf((float)h_imgII.ptr()[h_imgII.stride()*i+j] - (float)h_imgII_d.ptr()[h_imgII_d.stride()*i+j]) > 0.01f)
+                {
+                    bLoopVirgin = false;
+                }
+            }
+            else
+            {
+                ncvAssertPrintReturn(false, "Incorrect integral image test instance", false);
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+template <class T_in, class T_out>
+bool TestIntegralImage<T_in, T_out>::deinit()
+{
+    return true;
+}
+
+
+template class TestIntegralImage<Ncv8u, Ncv32u>;
+template class TestIntegralImage<Ncv32f, Ncv32f>;

modules/gpu/test/nvidia/TestIntegralImage.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testintegralimage_h_
+#define _testintegralimage_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+template <class T_in, class T_out>
+class TestIntegralImage : public NCVTestProvider
+{
+public:
+
+    TestIntegralImage(std::string testName, NCVTestSourceProvider<T_in> &src,
+                      Ncv32u width, Ncv32u height);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestIntegralImage(const TestIntegralImage&);
+	TestIntegralImage& operator=(const TestIntegralImage&);	
+
+    NCVTestSourceProvider<T_in> &src;
+    Ncv32u width;
+    Ncv32u height;
+};
+
+#endif // _testintegralimage_h_

modules/gpu/test/nvidia/TestIntegralImageSquared.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include "TestIntegralImageSquared.h"
+
+
+TestIntegralImageSquared::TestIntegralImageSquared(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                                                   Ncv32u width, Ncv32u height)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    width(width),
+    height(height)
+{
+}
+
+
+bool TestIntegralImageSquared::toString(std::ofstream &strOut)
+{
+    strOut << "width=" << width << std::endl;
+    strOut << "height=" << height << std::endl;
+    return true;
+}
+
+
+bool TestIntegralImageSquared::init()
+{
+    return true;
+}
+
+
+bool TestIntegralImageSquared::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32u widthSII = this->width + 1;
+    Ncv32u heightSII = this->height + 1;
+
+    NCVMatrixAlloc<Ncv8u> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv8u> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> d_imgSII(*this->allocatorGPU.get(), widthSII, heightSII);
+    ncvAssertReturn(d_imgSII.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> h_imgSII(*this->allocatorCPU.get(), widthSII, heightSII);
+    ncvAssertReturn(h_imgSII.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> h_imgSII_d(*this->allocatorCPU.get(), widthSII, heightSII);
+    ncvAssertReturn(h_imgSII_d.isMemAllocated(), false);
+
+    Ncv32u bufSize;
+    ncvStat = nppiStSqrIntegralGetSize_8u64u(NcvSize32u(this->width, this->height), &bufSize, this->devProp);
+    ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    NCVVectorAlloc<Ncv8u> d_tmpBuf(*this->allocatorGPU.get(), bufSize);
+    ncvAssertReturn(d_tmpBuf.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+
+    ncvAssertReturn(this->src.fill(h_img), false);
+
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStSqrIntegral_8u64u_C1R(d_img.ptr(), d_img.pitch(),
+                                          d_imgSII.ptr(), d_imgSII.pitch(),
+                                          NcvSize32u(this->width, this->height),
+                                          d_tmpBuf.ptr(), bufSize, this->devProp);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = d_imgSII.copySolid(h_imgSII_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStSqrIntegral_8u64u_C1R_host(h_img.ptr(), h_img.pitch(),
+                                               h_imgSII.ptr(), h_imgSII.pitch(),
+                                               NcvSize32u(this->width, this->height));
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    for (Ncv32u i=0; bLoopVirgin && i < h_img.height() + 1; i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < h_img.width() + 1; j++)
+        {
+            if (h_imgSII.ptr()[h_imgSII.stride()*i+j] != h_imgSII_d.ptr()[h_imgSII_d.stride()*i+j])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestIntegralImageSquared::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestIntegralImageSquared.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testintegralimagesquared_h_
+#define _testintegralimagesquared_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestIntegralImageSquared : public NCVTestProvider
+{
+public:
+
+    TestIntegralImageSquared(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                             Ncv32u width, Ncv32u height);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestIntegralImageSquared(const TestIntegralImageSquared&);
+	TestIntegralImageSquared& operator=(const TestIntegralImageSquared&);	
+
+    NCVTestSourceProvider<Ncv8u> &src;
+    Ncv32u width;
+    Ncv32u height;
+};
+
+#endif // _testintegralimagesquared_h_

modules/gpu/test/nvidia/TestRectStdDev.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include <math.h>
+
+#include "TestRectStdDev.h"
+
+
+TestRectStdDev::TestRectStdDev(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                               Ncv32u width, Ncv32u height, NcvRect32u rect, Ncv32f scaleFactor,
+                               NcvBool bTextureCache)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    width(width),
+    height(height),
+    rect(rect),
+    scaleFactor(scaleFactor),
+    bTextureCache(bTextureCache)
+{
+}
+
+
+bool TestRectStdDev::toString(std::ofstream &strOut)
+{
+    strOut << "width=" << width << std::endl;
+    strOut << "height=" << height << std::endl;
+    strOut << "rect=[" << rect.x << ", " << rect.y << ", " << rect.width << ", " << rect.height << "]\n";
+    strOut << "scaleFactor=" << scaleFactor << std::endl;
+    strOut << "bTextureCache=" << bTextureCache << std::endl;
+    return true;
+}
+
+
+bool TestRectStdDev::init()
+{
+    return true;
+}
+
+
+bool TestRectStdDev::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32s _normWidth = (Ncv32s)this->width - this->rect.x - this->rect.width + 1;
+    Ncv32s _normHeight = (Ncv32s)this->height - this->rect.y - this->rect.height + 1;
+    if (_normWidth <= 0 || _normHeight <= 0)
+    {
+        return true;
+    }
+    Ncv32u normWidth = (Ncv32u)_normWidth;
+    Ncv32u normHeight = (Ncv32u)_normHeight;
+    NcvSize32u szNormRoi(normWidth, normHeight);
+
+    Ncv32u widthII = this->width + 1;
+    Ncv32u heightII = this->height + 1;
+    Ncv32u widthSII = this->width + 1;
+    Ncv32u heightSII = this->height + 1;
+
+    NCVMatrixAlloc<Ncv8u> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv8u> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+
+    NCVMatrixAlloc<Ncv32u> d_imgII(*this->allocatorGPU.get(), widthII, heightII);
+    ncvAssertReturn(d_imgII.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32u> h_imgII(*this->allocatorCPU.get(), widthII, heightII);
+    ncvAssertReturn(h_imgII.isMemAllocated(), false);
+
+    NCVMatrixAlloc<Ncv64u> d_imgSII(*this->allocatorGPU.get(), widthSII, heightSII);
+    ncvAssertReturn(d_imgSII.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv64u> h_imgSII(*this->allocatorCPU.get(), widthSII, heightSII);
+    ncvAssertReturn(h_imgSII.isMemAllocated(), false);
+
+    NCVMatrixAlloc<Ncv32f> d_norm(*this->allocatorGPU.get(), normWidth, normHeight);
+    ncvAssertReturn(d_norm.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32f> h_norm(*this->allocatorCPU.get(), normWidth, normHeight);
+    ncvAssertReturn(h_norm.isMemAllocated(), false);
+    NCVMatrixAlloc<Ncv32f> h_norm_d(*this->allocatorCPU.get(), normWidth, normHeight);
+    ncvAssertReturn(h_norm_d.isMemAllocated(), false);
+
+    Ncv32u bufSizeII, bufSizeSII;
+    ncvStat = nppiStIntegralGetSize_8u32u(NcvSize32u(this->width, this->height), &bufSizeII, this->devProp);
+    ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    ncvStat = nppiStSqrIntegralGetSize_8u64u(NcvSize32u(this->width, this->height), &bufSizeSII, this->devProp);
+    ncvAssertReturn(NPPST_SUCCESS == ncvStat, false);
+    Ncv32u bufSize = bufSizeII > bufSizeSII ? bufSizeII : bufSizeSII;
+    NCVVectorAlloc<Ncv8u> d_tmpBuf(*this->allocatorGPU.get(), bufSize);
+    ncvAssertReturn(d_tmpBuf.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_img), false);
+
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStIntegral_8u32u_C1R(d_img.ptr(), d_img.pitch(),
+                                       d_imgII.ptr(), d_imgII.pitch(),
+                                       NcvSize32u(this->width, this->height),
+                                       d_tmpBuf.ptr(), bufSize, this->devProp);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStSqrIntegral_8u64u_C1R(d_img.ptr(), d_img.pitch(),
+                                          d_imgSII.ptr(), d_imgSII.pitch(),
+                                          NcvSize32u(this->width, this->height),
+                                          d_tmpBuf.ptr(), bufSize, this->devProp);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStRectStdDev_32f_C1R(d_imgII.ptr(), d_imgII.pitch(),
+                                       d_imgSII.ptr(), d_imgSII.pitch(),
+                                       d_norm.ptr(), d_norm.pitch(),
+                                       szNormRoi, this->rect,
+                                       this->scaleFactor,
+                                       this->bTextureCache);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = d_norm.copySolid(h_norm_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStIntegral_8u32u_C1R_host(h_img.ptr(), h_img.pitch(),
+                                          h_imgII.ptr(), h_imgII.pitch(),
+                                          NcvSize32u(this->width, this->height));
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStSqrIntegral_8u64u_C1R_host(h_img.ptr(), h_img.pitch(),
+                                             h_imgSII.ptr(), h_imgSII.pitch(),
+                                             NcvSize32u(this->width, this->height));
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    ncvStat = nppiStRectStdDev_32f_C1R_host(h_imgII.ptr(), h_imgII.pitch(),
+                                          h_imgSII.ptr(), h_imgSII.pitch(),
+                                          h_norm.ptr(), h_norm.pitch(),
+                                          szNormRoi, this->rect,
+                                          this->scaleFactor);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    const Ncv64f relEPS = 0.005;
+    for (Ncv32u i=0; bLoopVirgin && i < h_norm.height(); i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < h_norm.width(); j++)
+        {
+            Ncv64f absErr = fabs(h_norm.ptr()[h_norm.stride()*i+j] - h_norm_d.ptr()[h_norm_d.stride()*i+j]);
+            Ncv64f relErr = absErr / h_norm.ptr()[h_norm.stride()*i+j];
+
+            if (relErr > relEPS)
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+bool TestRectStdDev::deinit()
+{
+    return true;
+}

modules/gpu/test/nvidia/TestRectStdDev.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testrectstddev_h_
+#define _testrectstddev_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+class TestRectStdDev : public NCVTestProvider
+{
+public:
+
+    TestRectStdDev(std::string testName, NCVTestSourceProvider<Ncv8u> &src,
+                   Ncv32u width, Ncv32u height, NcvRect32u rect, Ncv32f scaleFactor,
+                   NcvBool bTextureCache);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestRectStdDev(const TestRectStdDev&);
+	TestRectStdDev& operator=(const TestRectStdDev&);	
+
+    NCVTestSourceProvider<Ncv8u> &src;
+    NcvRect32u rect;
+    Ncv32u width;
+    Ncv32u height;
+    Ncv32f scaleFactor;
+
+    NcvBool bTextureCache;
+};
+
+#endif // _testrectstddev_h_

modules/gpu/test/nvidia/TestResize.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include <math.h>
+
+#include "TestResize.h"
+
+
+template <class T>
+TestResize<T>::TestResize(std::string testName, NCVTestSourceProvider<T> &src,
+                          Ncv32u width, Ncv32u height, Ncv32u scaleFactor, NcvBool bTextureCache)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    width(width),
+    height(height),
+    scaleFactor(scaleFactor),
+    bTextureCache(bTextureCache)
+{
+}
+
+
+template <class T>
+bool TestResize<T>::toString(std::ofstream &strOut)
+{
+    strOut << "sizeof(T)=" << sizeof(T) << std::endl;
+    strOut << "width=" << width << std::endl;
+    strOut << "scaleFactor=" << scaleFactor << std::endl;
+    strOut << "bTextureCache=" << bTextureCache << std::endl;
+    return true;
+}
+
+
+template <class T>
+bool TestResize<T>::init()
+{
+    return true;
+}
+
+
+template <class T>
+bool TestResize<T>::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    Ncv32s smallWidth = this->width / this->scaleFactor;
+    Ncv32s smallHeight = this->height / this->scaleFactor;
+    if (smallWidth == 0 || smallHeight == 0)
+    {
+        return true;
+    }
+
+    NcvSize32u srcSize(this->width, this->height);
+
+    NCVMatrixAlloc<T> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+
+    NCVMatrixAlloc<T> d_small(*this->allocatorGPU.get(), smallWidth, smallHeight);
+    ncvAssertReturn(d_small.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_small(*this->allocatorCPU.get(), smallWidth, smallHeight);
+    ncvAssertReturn(h_small.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_small_d(*this->allocatorCPU.get(), smallWidth, smallHeight);
+    ncvAssertReturn(h_small_d.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_img), false);
+    NCV_SKIP_COND_END
+
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_BEGIN
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = nppiStDownsampleNearest_32u_C1R((Ncv32u *)d_img.ptr(), d_img.pitch(),
+                                                  (Ncv32u *)d_small.ptr(), d_small.pitch(),
+                                                  srcSize, this->scaleFactor,
+                                                  this->bTextureCache);
+    }
+    else if (sizeof(T) == sizeof(Ncv64u))
+    {
+        ncvStat = nppiStDownsampleNearest_64u_C1R((Ncv64u *)d_img.ptr(), d_img.pitch(),
+                                                  (Ncv64u *)d_small.ptr(), d_small.pitch(),
+                                                  srcSize, this->scaleFactor,
+                                                  this->bTextureCache);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect downsample test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+    ncvStat = d_small.copySolid(h_small_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    NCV_SKIP_COND_BEGIN
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = nppiStDownsampleNearest_32u_C1R_host((Ncv32u *)h_img.ptr(), h_img.pitch(),
+                                                       (Ncv32u *)h_small.ptr(), h_small.pitch(),
+                                                       srcSize, this->scaleFactor);
+    }
+    else if (sizeof(T) == sizeof(Ncv64u))
+    {
+        ncvStat = nppiStDownsampleNearest_64u_C1R_host((Ncv64u *)h_img.ptr(), h_img.pitch(),
+                                                       (Ncv64u *)h_small.ptr(), h_small.pitch(),
+                                                       srcSize, this->scaleFactor);
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect downsample test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    //const Ncv64f relEPS = 0.005;
+    for (Ncv32u i=0; bLoopVirgin && i < h_small.height(); i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < h_small.width(); j++)
+        {
+            if (h_small.ptr()[h_small.stride()*i+j] != h_small_d.ptr()[h_small_d.stride()*i+j])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+template <class T>
+bool TestResize<T>::deinit()
+{
+    return true;
+}
+
+
+template class TestResize<Ncv32u>;
+template class TestResize<Ncv64u>;

modules/gpu/test/nvidia/TestResize.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testresize_h_
+#define _testresize_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+template <class T>
+class TestResize : public NCVTestProvider
+{
+public:
+
+    TestResize(std::string testName, NCVTestSourceProvider<T> &src,
+               Ncv32u width, Ncv32u height, Ncv32u scaleFactor, NcvBool bTextureCache);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+	TestResize(const TestResize&);
+	TestResize& operator=(const TestResize&);	
+
+    NCVTestSourceProvider<T> &src;
+    Ncv32u width;
+    Ncv32u height;
+    Ncv32u scaleFactor;
+
+    NcvBool bTextureCache;
+};
+
+#endif // _testresize_h_

modules/gpu/test/nvidia/TestTranspose.cpp

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+
+#include <math.h>
+
+#include "TestTranspose.h"
+
+
+template <class T>
+TestTranspose<T>::TestTranspose(std::string testName, NCVTestSourceProvider<T> &src,
+                                Ncv32u width, Ncv32u height)
+    :
+    NCVTestProvider(testName),
+    src(src),
+    width(width),
+    height(height)
+{
+}
+
+
+template <class T>
+bool TestTranspose<T>::toString(std::ofstream &strOut)
+{
+    strOut << "sizeof(T)=" << sizeof(T) << std::endl;
+    strOut << "width=" << width << std::endl;
+    return true;
+}
+
+
+template <class T>
+bool TestTranspose<T>::init()
+{
+    return true;
+}
+
+
+template <class T>
+bool TestTranspose<T>::process()
+{
+    NCVStatus ncvStat;
+    bool rcode = false;
+
+    NcvSize32u srcSize(this->width, this->height);
+
+    NCVMatrixAlloc<T> d_img(*this->allocatorGPU.get(), this->width, this->height);
+    ncvAssertReturn(d_img.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_img(*this->allocatorCPU.get(), this->width, this->height);
+    ncvAssertReturn(h_img.isMemAllocated(), false);
+
+    NCVMatrixAlloc<T> d_dst(*this->allocatorGPU.get(), this->height, this->width);
+    ncvAssertReturn(d_dst.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_dst(*this->allocatorCPU.get(), this->height, this->width);
+    ncvAssertReturn(h_dst.isMemAllocated(), false);
+    NCVMatrixAlloc<T> h_dst_d(*this->allocatorCPU.get(), this->height, this->width);
+    ncvAssertReturn(h_dst_d.isMemAllocated(), false);
+
+    NCV_SET_SKIP_COND(this->allocatorGPU.get()->isCounting());
+    NCV_SKIP_COND_BEGIN
+    ncvAssertReturn(this->src.fill(h_img), false);
+    NCV_SKIP_COND_END
+
+    ncvStat = h_img.copySolid(d_img, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_BEGIN
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = nppiStTranspose_32u_C1R((Ncv32u *)d_img.ptr(), d_img.pitch(),
+                                          (Ncv32u *)d_dst.ptr(), d_dst.pitch(),
+                                          NcvSize32u(this->width, this->height));
+    }
+    else if (sizeof(T) == sizeof(Ncv64u))
+    {
+        ncvStat = nppiStTranspose_64u_C1R((Ncv64u *)d_img.ptr(), d_img.pitch(),
+                                        (Ncv64u *)d_dst.ptr(), d_dst.pitch(),
+                                        NcvSize32u(this->width, this->height));
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect transpose test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+    ncvStat = d_dst.copySolid(h_dst_d, 0);
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+
+    NCV_SKIP_COND_BEGIN
+    if (sizeof(T) == sizeof(Ncv32u))
+    {
+        ncvStat = nppiStTranspose_32u_C1R_host((Ncv32u *)h_img.ptr(), h_img.pitch(),
+                                               (Ncv32u *)h_dst.ptr(), h_dst.pitch(),
+                                               NcvSize32u(this->width, this->height));
+    }
+    else if (sizeof(T) == sizeof(Ncv64u))
+    {
+        ncvStat = nppiStTranspose_64u_C1R_host((Ncv64u *)h_img.ptr(), h_img.pitch(),
+                                               (Ncv64u *)h_dst.ptr(), h_dst.pitch(),
+                                               NcvSize32u(this->width, this->height));
+    }
+    else
+    {
+        ncvAssertPrintReturn(false, "Incorrect downsample test instance", false);
+    }
+    ncvAssertReturn(ncvStat == NPPST_SUCCESS, false);
+    NCV_SKIP_COND_END
+
+    //bit-to-bit check
+    bool bLoopVirgin = true;
+
+    NCV_SKIP_COND_BEGIN
+    //const Ncv64f relEPS = 0.005;
+    for (Ncv32u i=0; bLoopVirgin && i < this->width; i++)
+    {
+        for (Ncv32u j=0; bLoopVirgin && j < this->height; j++)
+        {
+            if (h_dst.ptr()[h_dst.stride()*i+j] != h_dst_d.ptr()[h_dst_d.stride()*i+j])
+            {
+                bLoopVirgin = false;
+            }
+        }
+    }
+    NCV_SKIP_COND_END
+
+    if (bLoopVirgin)
+    {
+        rcode = true;
+    }
+
+    return rcode;
+}
+
+
+template <class T>
+bool TestTranspose<T>::deinit()
+{
+    return true;
+}
+
+
+template class TestTranspose<Ncv32u>;
+template class TestTranspose<Ncv64u>;

modules/gpu/test/nvidia/TestTranspose.h

+/*
+ * Copyright 1993-2010 NVIDIA Corporation.  All rights reserved.
+ *
+ * NVIDIA Corporation and its licensors retain all intellectual 
+ * property and proprietary rights in and to this software and 
+ * related documentation and any modifications thereto.  
+ * Any use, reproduction, disclosure, or distribution of this 
+ * software and related documentation without an express license 
+ * agreement from NVIDIA Corporation is strictly prohibited.
+ */
+#ifndef _testtranspose_h_
+#define _testtranspose_h_
+
+#include "NCVTest.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+template <class T>
+class TestTranspose : public NCVTestProvider
+{
+public:
+
+    TestTranspose(std::string testName, NCVTestSourceProvider<T> &src,
+                  Ncv32u width, Ncv32u height);
+
+    virtual bool init();
+    virtual bool process();
+    virtual bool deinit();
+    virtual bool toString(std::ofstream &strOut);
+
+private:
+
+	TestTranspose(const TestTranspose&);
+	TestTranspose& operator=(const TestTranspose&);	
+
+    NCVTestSourceProvider<T> &src;
+    Ncv32u width;
+    Ncv32u height;
+};
+
+#endif // _testtranspose_h_

modules/gpu/test/nvidia/main_nvidia.cpp

+#pragma warning (disable : 4408 4201 4100)
+ 
+#include <cstdio>
+
+#include "NCV.hpp"
+#include "NCVHaarObjectDetection.hpp"
+
+#include "TestIntegralImage.h"
+#include "TestIntegralImageSquared.h"
+#include "TestRectStdDev.h"
+#include "TestResize.h"
+#include "TestCompact.h"
+#include "TestTranspose.h"
+
+#include "TestDrawRects.h"
+#include "TestHypothesesGrow.h"
+#include "TestHypothesesFilter.h"
+#include "TestHaarCascadeLoader.h"
+#include "TestHaarCascadeApplication.h"
+
+#include "NCVAutoTestLister.hpp"
+#include "NCVTestSourceProvider.hpp"
+
+
+template <class T_in, class T_out>
+void generateIntegralTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<T_in> &src,
+                           Ncv32u maxWidth, Ncv32u maxHeight)
+{
+    for (Ncv32f _i=1.0; _i<maxWidth; _i*=1.2f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "LinIntImgW%dH%d", i, 2);
+        testLister.add(new TestIntegralImage<T_in, T_out>(testName, src, i, 2));
+    }
+    for (Ncv32f _i=1.0; _i<maxHeight; _i*=1.2f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "LinIntImgW%dH%d", 2, i);
+        testLister.add(new TestIntegralImage<T_in, T_out>(testName, src, 2, i));
+    }
+
+    //test VGA
+    testLister.add(new TestIntegralImage<T_in, T_out>("LinIntImg_VGA", src, 640, 480));
+
+    //TODO: add tests of various resolutions up to 4096x4096
+}
+
+
+void generateSquaredIntegralTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv8u> &src,
+                                  Ncv32u maxWidth, Ncv32u maxHeight)
+{
+    for (Ncv32f _i=1.0; _i<maxWidth; _i*=1.2f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "SqIntImgW%dH%d", i, 32);
+        testLister.add(new TestIntegralImageSquared(testName, src, i, 32));
+    }
+    for (Ncv32f _i=1.0; _i<maxHeight; _i*=1.2f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "SqIntImgW%dH%d", 32, i);
+        testLister.add(new TestIntegralImageSquared(testName, src, 32, i));
+    }
+
+    //test VGA
+    testLister.add(new TestIntegralImageSquared("SqLinIntImg_VGA", src, 640, 480));
+
+    //TODO: add tests of various resolutions up to 4096x4096
+}
+
+
+void generateRectStdDevTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv8u> &src,
+                             Ncv32u maxWidth, Ncv32u maxHeight)
+{
+    NcvRect32u rect(1,1,18,18);
+
+    for (Ncv32f _i=32; _i<maxHeight/2 && _i < maxWidth/2; _i*=1.2f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "RectStdDevW%dH%d", i*2, i);
+        testLister.add(new TestRectStdDev(testName, src, i*2, i, rect, 1, true));
+        testLister.add(new TestRectStdDev(testName, src, i*2, i, rect, 1.5, false));
+        testLister.add(new TestRectStdDev(testName, src, i-1, i*2-1, rect, 1, false));
+        testLister.add(new TestRectStdDev(testName, src, i-1, i*2-1, rect, 2.5, true));
+    }
+
+    //test VGA
+    testLister.add(new TestRectStdDev("RectStdDev_VGA", src, 640, 480, rect, 1, true));
+
+    //TODO: add tests of various resolutions up to 4096x4096
+}
+
+
+template <class T>
+void generateResizeTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<T> &src)
+{
+    //test VGA
+    for (Ncv32u i=1; i<480; i+=3)
+    {
+        char testName[80];
+        sprintf(testName, "TestResize_VGA_s%d", i);
+        testLister.add(new TestResize<T>(testName, src, 640, 480, i, true));
+        testLister.add(new TestResize<T>(testName, src, 640, 480, i, false));
+    }
+
+    //test HD
+    for (Ncv32u i=1; i<1080; i+=5)
+    {
+        char testName[80];
+        sprintf(testName, "TestResize_1080_s%d", i);
+        testLister.add(new TestResize<T>(testName, src, 1920, 1080, i, true));
+        testLister.add(new TestResize<T>(testName, src, 1920, 1080, i, false));
+    }
+
+    //TODO: add tests of various resolutions up to 4096x4096
+}
+
+
+void generateNPPSTVectorTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv32u> &src, Ncv32u maxLength)
+{
+    //compaction
+    for (Ncv32f _i=256.0; _i<maxLength; _i*=1.1f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "Compaction%d", i);
+        testLister.add(new TestCompact(testName, src, i, 0xFFFFFFFF, 30));
+    }
+    for (Ncv32u i=1; i<260; i++)
+    {
+        char testName[80];
+        sprintf(testName, "Compaction%d", i);
+        testLister.add(new TestCompact(testName, src, i, 0xC001C0DE, 70));
+        testLister.add(new TestCompact(testName, src, i, 0xC001C0DE, 0));
+        testLister.add(new TestCompact(testName, src, i, 0xC001C0DE, 100));
+    }
+    for (Ncv32u i=256*256-256; i<256*256+257; i++)
+    {
+        char testName[80];
+        sprintf(testName, "Compaction%d", i);
+        testLister.add(new TestCompact(testName, src, i, 0xFFFFFFFF, 40));
+    }
+    for (Ncv32u i=256*256*256-10; i<256*256*256+10; i++)
+    {
+        char testName[80];
+        sprintf(testName, "Compaction%d", i);
+        testLister.add(new TestCompact(testName, src, i, 0x00000000, 2));
+    }
+}
+
+
+template <class T>
+void generateTransposeTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<T> &src)
+{
+    for (int i=2; i<64; i+=4)
+    {
+        for (int j=2; j<64; j+=4)
+        {
+            char testName[80];
+            sprintf(testName, "TestTranspose_%dx%d", i, j);
+            testLister.add(new TestTranspose<T>(testName, src, i, j));
+        }
+    }
+
+    for (int i=1; i<128; i+=1)
+    {
+        for (int j=1; j<2; j+=1)
+        {
+            char testName[80];
+            sprintf(testName, "TestTranspose_%dx%d", i, j);
+            testLister.add(new TestTranspose<T>(testName, src, i, j));
+        }
+    }
+
+    testLister.add(new TestTranspose<T>("TestTranspose_VGA", src, 640, 480));
+    testLister.add(new TestTranspose<T>("TestTranspose_HD1080", src, 1920, 1080));
+}
+
+
+template <class T>
+void generateDrawRectsTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<T> &src, NCVTestSourceProvider<Ncv32u> &src32u,
+                            Ncv32u maxWidth, Ncv32u maxHeight)
+{
+    for (Ncv32f _i=16.0; _i<maxWidth; _i*=1.1f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        Ncv32u j = maxHeight * i / maxWidth;
+        if (!j) continue;
+        char testName[80];
+        sprintf(testName, "DrawRectsW%dH%d", i, j);
+
+        if (sizeof(T) == sizeof(Ncv32u))
+        {
+            testLister.add(new TestDrawRects<T>(testName, src, src32u, i, j, i*j/1000+1, (T)0xFFFFFFFF));
+        }
+        else if (sizeof(T) == sizeof(Ncv8u))
+        {
+            testLister.add(new TestDrawRects<T>(testName, src, src32u, i, j, i*j/1000+1, (T)0xFF));
+        }
+        else
+        {
+            ncvAssertPrintCheck(false, "Attempted to instantiate non-existing DrawRects test suite");
+        }
+    }
+
+    //test VGA
+    testLister.add(new TestDrawRects<T>("DrawRects_VGA", src, src32u, 640, 480, 640*480/1000, (T)0xFF));
+
+    //TODO: add tests of various resolutions up to 4096x4096
+}
+
+
+void generateVectorTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv32u> &src, Ncv32u maxLength)
+{
+    //growth
+    for (Ncv32f _i=10.0; _i<maxLength; _i*=1.1f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "VectorGrow%d", i);
+        testLister.add(new TestHypothesesGrow(testName, src, 20, 20, 2.2f, i, i/2, i, i/4));
+        testLister.add(new TestHypothesesGrow(testName, src, 10, 42, 1.2f, i, i, i, 0));
+    }
+    testLister.add(new TestHypothesesGrow("VectorGrow01b", src, 10, 42, 1.2f, 10, 0, 10, 1));
+    testLister.add(new TestHypothesesGrow("VectorGrow11b", src, 10, 42, 1.2f, 10, 1, 10, 1));
+    testLister.add(new TestHypothesesGrow("VectorGrow10b", src, 10, 42, 1.2f, 10, 1, 10, 0));
+    testLister.add(new TestHypothesesGrow("VectorGrow00b", src, 10, 42, 1.2f, 10, 0, 10, 0));
+}
+
+
+void generateHypothesesFiltrationTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv32u> &src, Ncv32u maxLength)
+{
+    for (Ncv32f _i=1.0; _i<maxLength; _i*=1.1f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "HypFilter%d", i);
+        testLister.add(new TestHypothesesFilter(testName, src, i, 3, 0.2f));
+        testLister.add(new TestHypothesesFilter(testName, src, i, 0, 0.2f));
+        testLister.add(new TestHypothesesFilter(testName, src, i, 1, 0.1f));
+    }
+}
+
+
+void generateHaarLoaderTests(NCVAutoTestLister &testLister)
+{
+    testLister.add(new TestHaarCascadeLoader("haarcascade_eye.xml", "haarcascade_eye.xml"));
+    testLister.add(new TestHaarCascadeLoader("haarcascade_frontalface_alt.xml", "haarcascade_frontalface_alt.xml"));
+    testLister.add(new TestHaarCascadeLoader("haarcascade_frontalface_alt2.xml", "haarcascade_frontalface_alt2.xml"));
+    testLister.add(new TestHaarCascadeLoader("haarcascade_frontalface_alt_tree.xml", "haarcascade_frontalface_alt_tree.xml"));
+    testLister.add(new TestHaarCascadeLoader("haarcascade_eye_tree_eyeglasses.xml", "haarcascade_eye_tree_eyeglasses.xml"));
+}
+
+
+void generateHaarApplicationTests(NCVAutoTestLister &testLister, NCVTestSourceProvider<Ncv8u> &src,
+                                  Ncv32u maxWidth, Ncv32u maxHeight)
+{
+    for (Ncv32u i=20; i<512; i+=11)
+    {
+        for (Ncv32u j=20; j<128; j+=5)
+        {
+            char testName[80];
+            sprintf(testName, "HaarAppl%d_%d", i, j);
+            testLister.add(new TestHaarCascadeApplication(testName, src, "haarcascade_frontalface_alt.xml", j, i));
+        }
+    }
+    for (Ncv32f _i=20.0; _i<maxWidth; _i*=1.1f)
+    {
+        Ncv32u i = (Ncv32u)_i;
+        char testName[80];
+        sprintf(testName, "HaarAppl%d", i);
+        testLister.add(new TestHaarCascadeApplication(testName, src, "haarcascade_frontalface_alt.xml", i, i));
+    }
+}
+
+
+static void devNullOutput(const char *msg)
+{
+}
+
+
+bool main_nvidia()
+{
+    printf("Testing NVIDIA Computer Vision SDK\n");
+    printf("==================================\n");
+
+    ncvSetDebugOutputHandler(devNullOutput);
+
+    NCVAutoTestLister testListerII("NPPST Integral Image"                   );//,,true, false);
+    NCVAutoTestLister testListerSII("NPPST Squared Integral Image"          );//,,true, false);
+    NCVAutoTestLister testListerRStdDev("NPPST RectStdDev"                  );//,,true, false);
+    NCVAutoTestLister testListerResize("NPPST Resize"                       );//,,true, false);
+    NCVAutoTestLister testListerNPPSTVectorOperations("NPPST Vector Operations"  );//,,true, false);
+    NCVAutoTestLister testListerTranspose("NPPST Transpose"                 );//,,true, false);
+
+    NCVAutoTestLister testListerVectorOperations("Vector Operations"        );//,,true, false);
+    NCVAutoTestLister testListerHaarLoader("Haar Cascade Loader"            );//,,true, false);
+    NCVAutoTestLister testListerHaarAppl("Haar Cascade Application"         );//,,true, false);
+    NCVAutoTestLister testListerHypFiltration("Hypotheses Filtration"       );//,,true, false);
+    NCVAutoTestLister testListerVisualize("Visualization"                   );//,,true, false);
+
+    printf("Initializing data source providers\n");
+    NCVTestSourceProvider<Ncv32u> testSrcRandom_32u(2010, 0, 0xFFFFFFFF, 4096, 4096);
+    NCVTestSourceProvider<Ncv8u> testSrcRandom_8u(2010, 0, 255, 4096, 4096);
+    NCVTestSourceProvider<Ncv64u> testSrcRandom_64u(2010, 0, 0xFFFFFFFFFFFFFFFF, 4096, 4096);
+    NCVTestSourceProvider<Ncv8u> testSrcFacesVGA_8u("../../data/group_1_640x480_VGA.pgm");
+    NCVTestSourceProvider<Ncv32f> testSrcRandom_32f(2010, -1.0f, 1.0f, 4096, 4096);
+
+    printf("Generating NPPST test suites\n");
+    generateIntegralTests<Ncv8u, Ncv32u>(testListerII, testSrcRandom_8u, 4096, 4096);
+    generateIntegralTests<Ncv32f, Ncv32f>(testListerII, testSrcRandom_32f, 4096, 4096);
+    generateSquaredIntegralTests(testListerSII, testSrcRandom_8u, 4096, 4096);
+    generateRectStdDevTests(testListerRStdDev, testSrcRandom_8u, 4096, 4096);
+    generateResizeTests(testListerResize, testSrcRandom_32u);
+    generateResizeTests(testListerResize, testSrcRandom_64u);
+    generateNPPSTVectorTests(testListerNPPSTVectorOperations, testSrcRandom_32u, 4096*4096);
+    generateTransposeTests(testListerTranspose, testSrcRandom_32u);
+    generateTransposeTests(testListerTranspose, testSrcRandom_64u);
+
+    printf("Generating NCV test suites\n");
+    generateDrawRectsTests(testListerVisualize, testSrcRandom_8u, testSrcRandom_32u, 4096, 4096);
+    generateDrawRectsTests(testListerVisualize, testSrcRandom_32u, testSrcRandom_32u, 4096, 4096);
+    generateVectorTests(testListerVectorOperations, testSrcRandom_32u, 4096*4096);
+    generateHypothesesFiltrationTests(testListerHypFiltration, testSrcRandom_32u, 1024);
+    generateHaarLoaderTests(testListerHaarLoader);
+    generateHaarApplicationTests(testListerHaarAppl, testSrcFacesVGA_8u, 1280, 720);
+
+    // Indicate if at least one test failed
+    bool passed = true;
+
+    // Invoke all tests
+    passed &= testListerII.invoke();
+    passed &= testListerSII.invoke();
+    passed &= testListerRStdDev.invoke();
+    passed &= testListerResize.invoke();
+    passed &= testListerNPPSTVectorOperations.invoke();
+    passed &= testListerTranspose.invoke();
+    passed &= testListerVisualize.invoke();
+    passed &= testListerVectorOperations.invoke();
+    passed &= testListerHypFiltration.invoke();
+    passed &= testListerHaarLoader.invoke();
+    passed &= testListerHaarAppl.invoke();
+
+    return passed;
+}

modules/gpu/test/test_imgproc_gpu.cpp

@@ -907,7 +907,6 @@ TEST(copyMakeBorder, accuracy) { CV_GpuNppImageCopyMakeBorderTest test; test.saf
 TEST(warpAffine, accuracy) { CV_GpuNppImageWarpAffineTest test; test.safe_run(); }
 TEST(warpPerspective, accuracy) { CV_GpuNppImageWarpPerspectiveTest test; test.safe_run(); }
 TEST(integral, accuracy) { CV_GpuNppImageIntegralTest test; test.safe_run(); }
-//TEST(canny, accuracy) { CV_GpuNppImageCannyTest test; test.safe_run(); }
 TEST(cvtColor, accuracy) { CV_GpuCvtColorTest test; test.safe_run(); }
 TEST(histograms, accuracy) { CV_GpuHistogramsTest test; test.safe_run(); }
 TEST(cornerHearris, accuracy) { CV_GpuCornerHarrisTest test; test.safe_run(); }

modules/gpu/test/test_main.cpp

@@ -4,5 +4,3 @@ CV_TEST_MAIN("gpu")
 
 // Run test with --gtest_catch_exceptions flag to avoid runtime errors in 
 // the case when there is no GPU
-
-// TODO Add NVIDIA tests

modules/gpu/test/test_nvidia.cpp

+/*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"
+#include "cvconfig.h"
+
+class CV_NVidiaTestsCaller : public cvtest::BaseTest
+{
+public:
+    CV_NVidiaTestsCaller() {}
+    virtual ~CV_NVidiaTestsCaller() {}
+
+protected:
+    
+	void run( int )
+	{   		
+#if defined(HAVE_CUDA)
+		bool main_nvidia();
+
+		// Invoke all NVIDIA Staging tests and obtain the result
+		bool passed = main_nvidia();
+
+		if (passed)
+		    ts->set_failed_test_info(cvtest::TS::OK);
+		else
+		    ts->set_failed_test_info(cvtest::TS::FAIL_INVALID_OUTPUT);
+
+#else
+		ts->set_failed_test_info(cvtest::TS::SKIPPED);
+#endif
+	}   
+};
+
+TEST(NPP_Staging, DISABLED_multitest) { CV_NVidiaTestsCaller test; test.safe_run(); }