diff --git a/modules/core/include/opencv2/core/cuda.hpp b/modules/core/include/opencv2/core/cuda.hpp index a9c7a39a8f330a0f9e0dc8519f055f00a15c8b30..64bc53ef51dbe2517b422cdbbe3ed8f4dee6173a 100644 --- a/modules/core/include/opencv2/core/cuda.hpp +++ b/modules/core/include/opencv2/core/cuda.hpp @@ -528,6 +528,7 @@ public: private: Ptr impl_; + Event(const Ptr& impl); friend struct EventAccessor; }; diff --git a/modules/core/include/opencv2/core/cuda.inl.hpp b/modules/core/include/opencv2/core/cuda.inl.hpp index d9ab2ae4f3e2f0a3414da427c1ae90dd33256b12..01dc6d7c0b97fe3fda9307499333c0374ff77c55 100644 --- a/modules/core/include/opencv2/core/cuda.inl.hpp +++ b/modules/core/include/opencv2/core/cuda.inl.hpp @@ -540,6 +540,16 @@ Stream::Stream(const Ptr& impl) { } +//=================================================================================== +// Event +//=================================================================================== + +inline +Event::Event(const Ptr& impl) + : impl_(impl) +{ +} + //=================================================================================== // Initialization & Info //=================================================================================== diff --git a/modules/core/include/opencv2/core/cuda_stream_accessor.hpp b/modules/core/include/opencv2/core/cuda_stream_accessor.hpp index dd6589bcb67cc9a2bb0666c6975540a8505f489d..0f8ee9b2d4c01aedc82fe56c9684bb7890687a62 100644 --- a/modules/core/include/opencv2/core/cuda_stream_accessor.hpp +++ b/modules/core/include/opencv2/core/cuda_stream_accessor.hpp @@ -52,7 +52,7 @@ */ #include -#include "opencv2/core/cvdef.h" +#include "opencv2/core/cuda.hpp" namespace cv { @@ -62,14 +62,12 @@ namespace cv //! @addtogroup cudacore_struct //! @{ - class Stream; - class Event; - /** @brief Class that enables getting cudaStream_t from cuda::Stream */ struct StreamAccessor { CV_EXPORTS static cudaStream_t getStream(const Stream& stream); + CV_EXPORTS static Stream wrapStream(cudaStream_t stream); }; /** @brief Class that enables getting cudaEvent_t from cuda::Event @@ -77,6 +75,7 @@ namespace cv struct EventAccessor { CV_EXPORTS static cudaEvent_t getEvent(const Event& event); + CV_EXPORTS static Event wrapEvent(cudaEvent_t event); }; //! @} diff --git a/modules/core/src/cuda_stream.cpp b/modules/core/src/cuda_stream.cpp index d3b5545e94fee4c5a799b3a69b1b7d16b64cfac4..1ea8df37b9c752b430afd0be8052655e4e89a8e1 100644 --- a/modules/core/src/cuda_stream.cpp +++ b/modules/core/src/cuda_stream.cpp @@ -280,32 +280,37 @@ class cv::cuda::Stream::Impl { public: cudaStream_t stream; - Ptr stackAllocator_; + bool ownStream; + + Ptr stackAllocator; Impl(); - Impl(cudaStream_t stream); + explicit Impl(cudaStream_t stream); ~Impl(); }; -cv::cuda::Stream::Impl::Impl() : stream(0) +cv::cuda::Stream::Impl::Impl() : stream(0), ownStream(false) { cudaSafeCall( cudaStreamCreate(&stream) ); + ownStream = true; - stackAllocator_ = makePtr(stream); + stackAllocator = makePtr(stream); } -cv::cuda::Stream::Impl::Impl(cudaStream_t stream_) : stream(stream_) +cv::cuda::Stream::Impl::Impl(cudaStream_t stream_) : stream(stream_), ownStream(false) { - stackAllocator_ = makePtr(stream); + stackAllocator = makePtr(stream); } cv::cuda::Stream::Impl::~Impl() { - stackAllocator_.release(); + stackAllocator.release(); - if (stream) + if (stream && ownStream) + { cudaStreamDestroy(stream); + } } #endif @@ -516,6 +521,11 @@ cudaStream_t cv::cuda::StreamAccessor::getStream(const Stream& stream) return stream.impl_->stream; } +Stream cv::cuda::StreamAccessor::wrapStream(cudaStream_t stream) +{ + return Stream(makePtr(stream)); +} + #endif ///////////////////////////////////////////////////////////// @@ -660,7 +670,7 @@ void cv::cuda::setBufferPoolConfig(int deviceId, size_t stackSize, int stackCoun #ifdef HAVE_CUDA -cv::cuda::BufferPool::BufferPool(Stream& stream) : allocator_(stream.impl_->stackAllocator_.get()) +cv::cuda::BufferPool::BufferPool(Stream& stream) : allocator_(stream.impl_->stackAllocator.get()) { } @@ -693,20 +703,29 @@ class cv::cuda::Event::Impl { public: cudaEvent_t event; + bool ownEvent; - Impl(unsigned int flags); + explicit Impl(unsigned int flags); + explicit Impl(cudaEvent_t event); ~Impl(); }; -cv::cuda::Event::Impl::Impl(unsigned int flags) : event(0) +cv::cuda::Event::Impl::Impl(unsigned int flags) : event(0), ownEvent(false) { cudaSafeCall( cudaEventCreateWithFlags(&event, flags) ); + ownEvent = true; +} + +cv::cuda::Event::Impl::Impl(cudaEvent_t e) : event(e), ownEvent(false) +{ } cv::cuda::Event::Impl::~Impl() { - if (event) + if (event && ownEvent) + { cudaEventDestroy(event); + } } cudaEvent_t cv::cuda::EventAccessor::getEvent(const Event& event) @@ -714,6 +733,11 @@ cudaEvent_t cv::cuda::EventAccessor::getEvent(const Event& event) return event.impl_->event; } +Event cv::cuda::EventAccessor::wrapEvent(cudaEvent_t event) +{ + return Event(makePtr(event)); +} + #endif cv::cuda::Event::Event(CreateFlags flags) diff --git a/modules/cudaarithm/test/test_stream.cpp b/modules/cudaarithm/test/test_stream.cpp index c9a5e694f22212737a396efd74be417b1c7d8e34..785b10e7483fb6851e3232751eea19d732bb82f6 100644 --- a/modules/cudaarithm/test/test_stream.cpp +++ b/modules/cudaarithm/test/test_stream.cpp @@ -47,6 +47,7 @@ #include #include "opencv2/core/cuda.hpp" +#include "opencv2/core/cuda_stream_accessor.hpp" #include "opencv2/ts/cuda_test.hpp" using namespace cvtest; @@ -129,6 +130,27 @@ CUDA_TEST_P(Async, Convert) stream.waitForCompletion(); } +CUDA_TEST_P(Async, WrapStream) +{ + cudaStream_t cuda_stream = NULL; + ASSERT_EQ(cudaSuccess, cudaStreamCreate(&cuda_stream)); + + { + cv::cuda::Stream stream = cv::cuda::StreamAccessor::wrapStream(cuda_stream); + + d_src.upload(src, stream); + d_src.convertTo(d_dst, CV_32S, stream); + d_dst.download(dst, stream); + + Async* test = this; + stream.enqueueHostCallback(checkConvert, test); + + stream.waitForCompletion(); + } + + ASSERT_EQ(cudaSuccess, cudaStreamDestroy(cuda_stream)); +} + CUDA_TEST_P(Async, HostMemAllocator) { cv::cuda::Stream stream;