diff --git a/.bazelrc b/.bazelrc index af05ec14cf..1013f30f75 100644 --- a/.bazelrc +++ b/.bazelrc @@ -121,6 +121,6 @@ build --cxxopt=-DPYTHON_DISABLE=1 coverage --cxxopt=-DPYTHON_DISABLE=1 test --cxxopt=-DPYTHON_DISABLE=1 -build --cxxopt=-DOVMS_DUMP_TO_FILE=0 -coverage --cxxopt=-DOVMS_DUMP_TO_FILE=0 -test --cxxopt=-DOVMS_DUMP_TO_FILE=0 +build --cxxopt=-DOVMS_DUMP_TO_FILE=1 +coverage --cxxopt=-DOVMS_DUMP_TO_FILE=1 +test --cxxopt=-DOVMS_DUMP_TO_FILE=1 diff --git a/Dockerfile.openvino b/Dockerfile.openvino index 64dfebdc71..b4f4a06e6e 100644 --- a/Dockerfile.openvino +++ b/Dockerfile.openvino @@ -170,6 +170,26 @@ RUN cd flatbuffers && mkdir build && cd build && cmake -G "Unix Makefiles" -DCMA ## End of base image +## GPU OCL support +ARG INSTALL_DRIVER_VERSION="23.13.26032" +ARG GPU=0 +if [ "$GPU" == "1" ] ; then \ + apt-get update && apt-get install -y ocl-icd-opencl-dev libnuma1 ocl-icd-libopencl1 --no-install-recommends && rm -rf /var/lib/apt/lists/* && \ + case $INSTALL_DRIVER_VERSION in \ + "23.22.26516") \ + mkdir /tmp/gpu_deps && cd /tmp/gpu_deps ; \ + curl -L -O https://github.com/intel/compute-runtime/releases/download/23.22.26516.18/intel-level-zero-gpu_1.3.26516.18_amd64.deb ; \ + curl -L -O https://github.com/intel/intel-graphics-compiler/releases/download/igc-1.0.14062.11/intel-igc-core_1.0.14062.11_amd64.deb ; \ + curl -L -O https://github.com/intel/intel-graphics-compiler/releases/download/igc-1.0.14062.11/intel-igc-opencl_1.0.14062.11_amd64.deb ; \ + curl -L -O https://github.com/intel/compute-runtime/releases/download/23.22.26516.18/intel-opencl-icd_23.22.26516.18_amd64.deb ; \ + curl -L -O https://github.com/intel/compute-runtime/releases/download/23.22.26516.18/libigdgmm12_22.3.0_amd64.deb ; \ + dpkg -i *.deb && rm -Rf /tmp/gpu_deps ; \ + ;; \ + esac ; \ +fi ; \ +apt-get clean ; \ +rm -rf /var/lib/apt/lists/* && rm -rf /tmp/* ; + FROM base as build RUN wget -O video.mp4 "https://www.pexels.com/download/video/3044127/?fps=24.0&h=1080&w=1920" diff --git a/Makefile b/Makefile index 1fc4abf41b..8b8084cbba 100644 --- a/Makefile +++ b/Makefile @@ -23,6 +23,8 @@ INPUT_VIDEO_LINK ?= "https://www.pexels.com/download/video/3044127/?fps=24.0&h=1 OVMS_COMMIT ?="7bf5d7fc217a81ce4b0a089c7679ecb77e6c544c" JOBS ?= $(shell python3 -c 'import multiprocessing as mp; print(mp.cpu_count())') DLDT_PACKAGE_URL ?= https://storage.openvinotoolkit.org/repositories/openvino/packages/2024.0/linux/l_openvino_toolkit_ubuntu20_2024.0.0.14509.34caeefd078_x86_64.tgz +INSTALL_DRIVER_VERSION ?= "23.22.26516" +GPU ?= 0 # Targets to use outside running mediapipe_ovms container docker_build: @@ -32,6 +34,8 @@ docker_build: --build-arg JOBS=$(JOBS) . \ --build-arg OVMS_COMMIT=$(OVMS_COMMIT) \ -t $(OVMS_MEDIA_DOCKER_IMAGE):$(OVMS_MEDIA_IMAGE_TAG) + --build-arg INSTALL_DRIVER_VERSION=$(INSTALL_DRIVER_VERSION)\ + --build-arg GPU=$(GPU)\ tests: run_unit_tests run_hello_world run_hello_ovms run_hello_ovms: diff --git a/WORKSPACE b/WORKSPACE index 3bcaa674db..3316d30e3c 100644 --- a/WORKSPACE +++ b/WORKSPACE @@ -825,6 +825,12 @@ new_local_repository( path = "/opt/intel/openvino/runtime", ) +new_git_repository( + name = "openvino", + remote = "https:///github.com/openvinotoolkit/openvino/", + commit = "c3c409ee133ffb26bf8fd5570ef50a7c004839a4" +) + git_repository( name = "oneTBB", branch = "v2021.10.0", diff --git a/mediapipe/calculators/image/image_transformation_calculator.cc b/mediapipe/calculators/image/image_transformation_calculator.cc index dbf8f73374..6c9be60978 100644 --- a/mediapipe/calculators/image/image_transformation_calculator.cc +++ b/mediapipe/calculators/image/image_transformation_calculator.cc @@ -190,6 +190,7 @@ class ImageTransformationCalculator : public CalculatorBase { private: absl::Status RenderCpu(CalculatorContext* cc); absl::Status RenderGpu(CalculatorContext* cc); + absl::Status RenderOpenCl(CalculatorContext* cc); absl::Status GlSetup(); void ComputeOutputDimensions(int input_width, int input_height, @@ -411,7 +412,8 @@ absl::Status ImageTransformationCalculator::Process(CalculatorContext* cc) { if (cc->Inputs().Tag(kImageFrameTag).IsEmpty()) { return absl::OkStatus(); } - return RenderCpu(cc); + //return RenderCpu(cc); + return RenderOpenCl(cc); } return absl::OkStatus(); } @@ -442,6 +444,137 @@ absl::Status ImageTransformationCalculator::Close(CalculatorContext* cc) { return absl::OkStatus(); } +absl::Status ImageTransformationCalculator::RenderOpenCl(CalculatorContext* cc) { + // TODO UMAT + cv::UMatUsageFlags usageFlags = cv::USAGE_ALLOCATE_HOST_MEMORY; + cv::UMat input_mat = cv::UMat(usageFlags); + mediapipe::ImageFormat::Format format; + + const auto& input = cc->Inputs().Tag(kImageFrameTag).Get(); + //std::cout << " cc->Inputs().Tag(kImageFrameTag).Get(); " << std::endl; + input_mat = formats::MatView(const_cast(&input), usageFlags); + //input_mat = formats::MatView(&input, cv::USAGE_ALLOCATE_SHARED_MEMORY); + std::string fileName = std::string("./imageTrans/input") + std::to_string(cc->InputTimestamp().Value()); + //dumpMatToFile(fileName, input_mat); + + format = input.Format(); + + const int input_width = input_mat.cols; + const int input_height = input_mat.rows; + int output_width; + int output_height; + ComputeOutputDimensions(input_width, input_height, &output_width, + &output_height); + + if (output_width_ > 0 && output_height_ > 0) { + // TODO umat + cv::UMat scaled_mat = cv::UMat(usageFlags); + if (scale_mode_ == mediapipe::ScaleMode_Mode_STRETCH) { + int scale_flag = + input_mat.cols > output_width_ && input_mat.rows > output_height_ + ? cv::INTER_AREA + : cv::INTER_LINEAR; + cv::resize(input_mat, scaled_mat, cv::Size(output_width_, output_height_), + 0, 0, scale_flag); + // DEFAULT + // std::cout << "SCALE - STRETCH" <(output_width_) / input_width, + static_cast(output_height_) / input_height); + const int target_width = std::round(input_width * scale); + const int target_height = std::round(input_height * scale); + int scale_flag = scale < 1.0f ? cv::INTER_AREA : cv::INTER_LINEAR; + if (scale_mode_ == mediapipe::ScaleMode_Mode_FIT) { + cv::UMat intermediate_mat = cv::UMat(usageFlags); + cv::resize(input_mat, intermediate_mat, + cv::Size(target_width, target_height), 0, 0, scale_flag); + const int top = (output_height_ - target_height) / 2; + const int bottom = output_height_ - target_height - top; + const int left = (output_width_ - target_width) / 2; + const int right = output_width_ - target_width - left; + cv::copyMakeBorder(intermediate_mat, scaled_mat, top, bottom, left, + right, + options_.constant_padding() ? cv::BORDER_CONSTANT + : cv::BORDER_REPLICATE, + padding_color_); + } else { + cv::resize(input_mat, scaled_mat, cv::Size(target_width, target_height), + 0, 0, scale_flag); + output_width = target_width; + output_height = target_height; + } + } + input_mat = scaled_mat; + } + + if (cc->Outputs().HasTag("LETTERBOX_PADDING")) { + auto padding = absl::make_unique>(); + ComputeOutputLetterboxPadding(input_width, input_height, output_width, + output_height, padding.get()); + cc->Outputs() + .Tag("LETTERBOX_PADDING") + .Add(padding.release(), cc->InputTimestamp()); + } + + //TODO umat + cv::UMat rotated_mat = cv::UMat(usageFlags); + cv::Size rotated_size(output_width, output_height); + if (input_mat.size() == rotated_size) { + const int angle = RotationModeToDegrees(rotation_); + cv::Point2f src_center(input_mat.cols / 2.0, input_mat.rows / 2.0); + // TODO UMAT + cv::Mat rotation_mat = cv::getRotationMatrix2D(src_center, angle, 1.0); + cv::warpAffine(input_mat, rotated_mat, rotation_mat, rotated_size); + // DEFAULT + // std::cout << "ROTATE - MATRIX" < output_frame( + new ImageFrame(flipped_mat, format, output_width, output_height)); + + fileName = std::string("./imageTrans/output") + std::to_string(cc->InputTimestamp().Value()); + //dumpMatToFile(fileName, flipped_mat); + + cc->Outputs() + .Tag(kImageFrameTag) + .Add(output_frame.release(), cc->InputTimestamp()); + + return absl::OkStatus(); +} + absl::Status ImageTransformationCalculator::RenderCpu(CalculatorContext* cc) { cv::Mat input_mat; mediapipe::ImageFormat::Format format; diff --git a/mediapipe/calculators/util/annotation_overlay_calculator.cc b/mediapipe/calculators/util/annotation_overlay_calculator.cc index 5afede99d2..5f6571d02a 100644 --- a/mediapipe/calculators/util/annotation_overlay_calculator.cc +++ b/mediapipe/calculators/util/annotation_overlay_calculator.cc @@ -143,6 +143,9 @@ class AnnotationOverlayCalculator : public CalculatorBase { absl::Status RenderToCpu(CalculatorContext* cc, const ImageFormat::Format& target_format, uchar* data_image); + absl::Status RenderToOpencl( + CalculatorContext* cc, const ImageFormat::Format& target_format, + cv::Mat& input_mat); absl::Status GlRender(CalculatorContext* cc); template @@ -345,8 +348,10 @@ absl::Status AnnotationOverlayCalculator::Process(CalculatorContext* cc) { #endif // !MEDIAPIPE_DISABLE_GPU } else { // Copy the rendered image to output. - uchar* image_mat_ptr = image_mat->data; - MP_RETURN_IF_ERROR(RenderToCpu(cc, target_format, image_mat_ptr)); + //uchar* image_mat_ptr = image_mat->data; + //MP_RETURN_IF_ERROR(RenderToCpu(cc, target_format, image_mat_ptr)); + + MP_RETURN_IF_ERROR(RenderToOpencl(cc, target_format, *image_mat.get())); } return absl::OkStatus(); @@ -365,6 +370,25 @@ absl::Status AnnotationOverlayCalculator::Close(CalculatorContext* cc) { return absl::OkStatus(); } +absl::Status AnnotationOverlayCalculator::RenderToOpencl( + CalculatorContext* cc, const ImageFormat::Format& target_format, + cv::Mat& input_mat) { + // TODO switch to umat + cv::UMat copy; + input_mat.copyTo(copy); + + std::unique_ptr output_frame( + new ImageFrame(copy, target_format, renderer_->GetImageWidth(), renderer_->GetImageHeight())); + + if (cc->Outputs().HasTag(kImageFrameTag)) { + cc->Outputs() + .Tag(kImageFrameTag) + .Add(output_frame.release(), cc->InputTimestamp()); + } + + return absl::OkStatus(); +} + absl::Status AnnotationOverlayCalculator::RenderToCpu( CalculatorContext* cc, const ImageFormat::Format& target_format, uchar* data_image) { diff --git a/mediapipe/examples/desktop/demo_run_graph_main.cc b/mediapipe/examples/desktop/demo_run_graph_main.cc index 3b6465eec6..fd261f6f38 100644 --- a/mediapipe/examples/desktop/demo_run_graph_main.cc +++ b/mediapipe/examples/desktop/demo_run_graph_main.cc @@ -16,6 +16,8 @@ #include #include +#include + #include "absl/flags/flag.h" #include "absl/flags/parse.h" #include "mediapipe/framework/calculator_framework.h" @@ -86,8 +88,14 @@ absl::Status RunMPPGraph() { LOG(INFO) << "Start grabbing and processing frames."; bool grab_frames = true; int count_frames = 0; - auto begin = std::chrono::high_resolution_clock::now(); + + OpenClWrapper ocl; + ocl.initOpenCL(); + + int max_frame = 0; + auto begin = std::chrono::high_resolution_clock::now(); + cv::UMatUsageFlags usageFlags = cv::USAGE_ALLOCATE_HOST_MEMORY; while (grab_frames) { // Capture opencv camera or video frame. cv::Mat camera_frame_raw; @@ -101,20 +109,29 @@ absl::Status RunMPPGraph() { break; } count_frames+=1; - cv::Mat camera_frame; - cv::cvtColor(camera_frame_raw, camera_frame, cv::COLOR_BGR2RGB); + cv::UMat camera_frame = cv::UMat(usageFlags); + // = cv::UMat(camera_frame_raw.rows, camera_frame_raw.cols, camera_frame_raw.type ,cv::USAGE_ALLOCATE_SHARED_MEMORY) + // SEGFAULT icv_k0_ownsCopy_8u_repE9 cv::cvtColor(camera_frame_raw, camera_frame, cv::COLOR_BGR2RGB); + //cv::cvtColor(camera_frame_raw, camera_frame, cv::COLOR_BGR2RGBA); + camera_frame_raw.copyTo(camera_frame); if (!load_video) { cv::flip(camera_frame, camera_frame, /*flipcode=HORIZONTAL*/ 1); } // Wrap Mat into an ImageFrame. auto input_frame = absl::make_unique( + camera_frame, mediapipe::ImageFormat::SRGB, camera_frame.cols, camera_frame.rows, mediapipe::ImageFrame::kDefaultAlignmentBoundary); - cv::Mat input_frame_mat = mediapipe::formats::MatView(input_frame.get()); - camera_frame.copyTo(input_frame_mat); - // Send image packet into the graph. + // mediapipe::ImageFormat::SRGBA, camera_frame.cols, camera_frame.rows, + //cv::UMat input_frame_mat = mediapipe::formats::MatView(input_frame.get(), cv::USAGE_ALLOCATE_SHARED_MEMORY); + //camera_frame.copyTo(input_frame_mat); + std::string fileName = std::string("./imageDemo/input") + std::to_string(count_frames); + //dumpMatToFile(fileName, camera_frame); + + + // Send image packet INFO the graph. size_t frame_timestamp_us = (double)cv::getTickCount() / (double)cv::getTickFrequency() * 1e6; MP_RETURN_IF_ERROR(graph.AddPacketToInputStream( @@ -127,8 +144,8 @@ absl::Status RunMPPGraph() { auto& output_frame = packet.Get(); // Convert back to opencv for display or saving. - cv::Mat output_frame_mat = mediapipe::formats::MatView(&output_frame); - cv::cvtColor(output_frame_mat, output_frame_mat, cv::COLOR_RGB2BGR); + cv::UMat output_frame_mat = mediapipe::formats::MatView(const_cast(&output_frame), usageFlags); + // SEGFAULT icv_k0_ownsCopy_8u_repE9 cv::cvtColor(output_frame_mat, output_frame_mat, cv::COLOR_RGB2BGR); if (save_video) { if (!writer.isOpened()) { LOG(INFO) << "Prepare video writer."; @@ -144,6 +161,10 @@ absl::Status RunMPPGraph() { const int pressed_key = cv::waitKey(5); if (pressed_key >= 0 && pressed_key != 255) grab_frames = false; } + + if (count_frames == max_frame) { + break; + } } auto duration = std::chrono::duration_cast(std::chrono::high_resolution_clock::now() - begin); auto totalTime = duration.count(); diff --git a/mediapipe/examples/desktop/demo_run_graph_main_gpu.cc b/mediapipe/examples/desktop/demo_run_graph_main_gpu.cc index 8336e56705..ecfb9ba5fd 100644 --- a/mediapipe/examples/desktop/demo_run_graph_main_gpu.cc +++ b/mediapipe/examples/desktop/demo_run_graph_main_gpu.cc @@ -94,8 +94,11 @@ absl::Status RunMPPGraph() { LOG(INFO) << "Start grabbing and processing frames."; bool grab_frames = true; + auto begin = std::chrono::high_resolution_clock::now(); + int count_frames = 0; while (grab_frames) { // Capture opencv camera or video frame. + count_frames+=1; cv::Mat camera_frame_raw; capture >> camera_frame_raw; if (camera_frame_raw.empty()) { @@ -183,7 +186,13 @@ absl::Status RunMPPGraph() { if (pressed_key >= 0 && pressed_key != 255) grab_frames = false; } } + auto duration = std::chrono::duration_cast(std::chrono::high_resolution_clock::now() - begin); + auto totalTime = duration.count(); + float avgFps = (1000000 * (float)(count_frames) / (float)totalTime); + float avgLatencyms = 1000 / avgFps; + LOG(INFO) << "Frames:" << count_frames << ", Duration [ms]:" << totalTime / 1000 << ", FPS:" << avgFps << ", Avg latency [ms]:" << avgLatencyms; + LOG(INFO) << "Shutting down."; if (writer.isOpened()) writer.release(); MP_RETURN_IF_ERROR(graph.CloseInputStream(kInputStream)); diff --git a/mediapipe/examples/desktop/face_detection/BUILD b/mediapipe/examples/desktop/face_detection/BUILD index cf647e0de6..1b21398014 100644 --- a/mediapipe/examples/desktop/face_detection/BUILD +++ b/mediapipe/examples/desktop/face_detection/BUILD @@ -40,6 +40,7 @@ cc_binary( name = "face_detection_gpu", data = ["//mediapipe/modules/face_detection:face_detection_short_range.tflite"], deps = [ + "@ovms//src:ovms_lib", "//mediapipe/examples/desktop:demo_run_graph_main_gpu", "//mediapipe/graphs/face_detection:desktop_live_gpu_calculators", ], diff --git a/mediapipe/framework/formats/BUILD b/mediapipe/framework/formats/BUILD index b23209f7d4..a4590f9d9c 100644 --- a/mediapipe/framework/formats/BUILD +++ b/mediapipe/framework/formats/BUILD @@ -147,8 +147,10 @@ cc_library( cc_library( name = "image_frame", - srcs = ["image_frame.cc"], - hdrs = ["image_frame.h"], + srcs = ["image_frame.cc", + "helpers.cpp"], + hdrs = ["image_frame.h", + "helpers.hpp"], deps = [ ":image_format_cc_proto", "//mediapipe/framework:port", @@ -162,11 +164,13 @@ cc_library( "@com_google_absl//absl/base:core_headers", "@com_google_absl//absl/memory", "@com_google_absl//absl/strings", + "//mediapipe/framework/port:opencv_core", ] + select({ "//conditions:default": [ ], "//mediapipe/framework:disable_rtti_and_exceptions": [], }), + linkopts = ["-lOpenCL"], ) cc_library( diff --git a/mediapipe/framework/formats/helpers.cpp b/mediapipe/framework/formats/helpers.cpp new file mode 100644 index 0000000000..f6acecd2f1 --- /dev/null +++ b/mediapipe/framework/formats/helpers.cpp @@ -0,0 +1,686 @@ +/* +// The example of interoperability between OpenCL and OpenCV. +// This will loop through frames of video either from input media file +// or camera device and do processing of these data in OpenCL and then +// in OpenCV. In OpenCL it does inversion of pixels in left half of frame and +// in OpenCV it does blurring in the right half of frame. +*/ +#include +#include +#include +#include +#include +#include +#include +#include +#include + +#define CL_USE_DEPRECATED_OPENCL_1_1_APIS +#define CL_USE_DEPRECATED_OPENCL_1_2_APIS +#define CL_USE_DEPRECATED_OPENCL_2_0_APIS // eliminate build warning +#define CL_TARGET_OPENCL_VERSION 200 // 2.0 + +#ifdef __APPLE__ +#define CL_SILENCE_DEPRECATION +#include +#else +#include +#endif +#include "mediapipe/framework/formats/helpers.hpp" + +#include +#include +#include +#include +#include + + +using namespace std; +using namespace cv; +static std::mutex g_dump_mutex; +static std::mutex g_dump_mutex2; + +void dumpMatToFile(std::string& fileName, cv::UMat& umat){ + std::lock_guard guard(g_dump_mutex); + cv::FileStorage file(fileName, cv::FileStorage::WRITE); + cv::Mat input; + //file << "camera_frame_raw" << camera_frame_raw; + umat.copyTo(input); + file << "input_frame_mat" << input; + + file.release(); +} + +void dumpMatToFile(std::string& fileName, cv::Mat& umat){ + std::lock_guard guard(g_dump_mutex2); + cv::FileStorage file(fileName, cv::FileStorage::WRITE); + file << "input_frame_mat" << umat; + + file.release(); +} + +cl_channel_order GetChannelOrderFromMatType(int cvMatType) { + switch (cvMatType) { + case CV_8UC1: + case CV_8SC1: + std::cout<< "Using CL_R " << endl; + return CL_R; + case CV_8UC2: + case CV_8SC2: + std::cout<< "Using CL_RG " << endl; + return CL_RG; + case CV_8UC3: + case CV_8SC3: + std::cout<< "Using CL_RGB " << endl; + return CL_RGB; + case CV_8UC4: + case CV_8SC4: + std::cout<< "Using CL_RGBA " << endl; + return CL_RGBA; + default: + { + std::cout<< "Using default CL_R for CV type: " << cvMatType << endl; + return CL_R; // Default to single channel + } + } +} + + +cl_channel_type GetChannelDataTypeFromOrder(cl_channel_order cl_order) { + switch (cl_order) { + case CL_R: + case CL_A: + return CL_UNSIGNED_INT8; // Or your desired data type + std::cout<< "Using CL_UNSIGNED_INT8 " << endl; + case CL_RG: + return CL_UNSIGNED_INT16; // Or your desired data type + case CL_RGB: + std::cout<< "Using CL_HALF_FLOAT " << endl; + return CL_HALF_FLOAT; // Or your desired data type + case CL_RGBA: + std::cout<< "Using CL_FLOAT " << endl; + return CL_FLOAT; // Or your desired data type + // Add more cases for other channel orders as needed + default: + // Handle unsupported channel order + std::cout<< "Using default CL_UNSIGNED_INT8 for channel order: " << cl_order << endl; + return CL_UNSIGNED_INT8; + break; + } +} + +const char* clGetErrorString(int errorCode) { + switch (errorCode) { + case 0: return "CL_SUCCESS"; + case -1: return "CL_DEVICE_NOT_FOUND"; + case -2: return "CL_DEVICE_NOT_AVAILABLE"; + case -3: return "CL_COMPILER_NOT_AVAILABLE"; + case -4: return "CL_MEM_OBJECT_ALLOCATION_FAILURE"; + case -5: return "CL_OUT_OF_RESOURCES"; + case -6: return "CL_OUT_OF_HOST_MEMORY"; + case -7: return "CL_PROFILING_INFO_NOT_AVAILABLE"; + case -8: return "CL_MEM_COPY_OVERLAP"; + case -9: return "CL_IMAGE_FORMAT_MISMATCH"; + case -10: return "CL_IMAGE_FORMAT_NOT_SUPPORTED"; + case -12: return "CL_MAP_FAILURE"; + case -13: return "CL_MISALIGNED_SUB_BUFFER_OFFSET"; + case -14: return "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST"; + case -15: return "CL_COMPILE_PROGRAM_FAILURE"; + case -16: return "CL_LINKER_NOT_AVAILABLE"; + case -17: return "CL_LINK_PROGRAM_FAILURE"; + case -18: return "CL_DEVICE_PARTITION_FAILED"; + case -19: return "CL_KERNEL_ARG_INFO_NOT_AVAILABLE"; + case -30: return "CL_INVALID_VALUE"; + case -31: return "CL_INVALID_DEVICE_TYPE"; + case -32: return "CL_INVALID_PLATFORM"; + case -33: return "CL_INVALID_DEVICE"; + case -34: return "CL_INVALID_CONTEXT"; + case -35: return "CL_INVALID_QUEUE_PROPERTIES"; + case -36: return "CL_INVALID_COMMAND_QUEUE"; + case -37: return "CL_INVALID_HOST_PTR"; + case -38: return "CL_INVALID_MEM_OBJECT"; + case -39: return "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR"; + case -40: return "CL_INVALID_IMAGE_SIZE"; + case -41: return "CL_INVALID_SAMPLER"; + case -42: return "CL_INVALID_BINARY"; + case -43: return "CL_INVALID_BUILD_OPTIONS"; + case -44: return "CL_INVALID_PROGRAM"; + case -45: return "CL_INVALID_PROGRAM_EXECUTABLE"; + case -46: return "CL_INVALID_KERNEL_NAME"; + case -47: return "CL_INVALID_KERNEL_DEFINITION"; + case -48: return "CL_INVALID_KERNEL"; + case -49: return "CL_INVALID_ARG_INDEX"; + case -50: return "CL_INVALID_ARG_VALUE"; + case -51: return "CL_INVALID_ARG_SIZE"; + case -52: return "CL_INVALID_KERNEL_ARGS"; + case -53: return "CL_INVALID_WORK_DIMENSION"; + case -54: return "CL_INVALID_WORK_GROUP_SIZE"; + case -55: return "CL_INVALID_WORK_ITEM_SIZE"; + case -56: return "CL_INVALID_GLOBAL_OFFSET"; + case -57: return "CL_INVALID_EVENT_WAIT_LIST"; + case -58: return "CL_INVALID_EVENT"; + case -59: return "CL_INVALID_OPERATION"; + case -60: return "CL_INVALID_GL_OBJECT"; + case -61: return "CL_INVALID_BUFFER_SIZE"; + case -62: return "CL_INVALID_MIP_LEVEL"; + case -63: return "CL_INVALID_GLOBAL_WORK_SIZE"; + case -64: return "CL_INVALID_PROPERTY"; + case -65: return "CL_INVALID_IMAGE_DESCRIPTOR"; + case -66: return "CL_INVALID_COMPILER_OPTIONS"; + case -67: return "CL_INVALID_LINKER_OPTIONS"; + case -68: return "CL_INVALID_DEVICE_PARTITION_COUNT"; + case -69: return "CL_INVALID_PIPE_SIZE"; + case -70: return "CL_INVALID_DEVICE_QUEUE"; + case -71: return "CL_INVALID_SPEC_ID"; + case -72: return "CL_MAX_SIZE_RESTRICTION_EXCEEDED"; + case -1002: return "CL_INVALID_D3D10_DEVICE_KHR"; + case -1003: return "CL_INVALID_D3D10_RESOURCE_KHR"; + case -1004: return "CL_D3D10_RESOURCE_ALREADY_ACQUIRED_KHR"; + case -1005: return "CL_D3D10_RESOURCE_NOT_ACQUIRED_KHR"; + case -1006: return "CL_INVALID_D3D11_DEVICE_KHR"; + case -1007: return "CL_INVALID_D3D11_RESOURCE_KHR"; + case -1008: return "CL_D3D11_RESOURCE_ALREADY_ACQUIRED_KHR"; + case -1009: return "CL_D3D11_RESOURCE_NOT_ACQUIRED_KHR"; + case -1010: return "CL_INVALID_DX9_MEDIA_ADAPTER_KHR"; + case -1011: return "CL_INVALID_DX9_MEDIA_SURFACE_KHR"; + case -1012: return "CL_DX9_MEDIA_SURFACE_ALREADY_ACQUIRED_KHR"; + case -1013: return "CL_DX9_MEDIA_SURFACE_NOT_ACQUIRED_KHR"; + case -1093: return "CL_INVALID_EGL_OBJECT_KHR"; + case -1092: return "CL_EGL_RESOURCE_NOT_ACQUIRED_KHR"; + case -1001: return "CL_PLATFORM_NOT_FOUND_KHR"; + case -1057: return "CL_DEVICE_PARTITION_FAILED_EXT"; + case -1058: return "CL_INVALID_PARTITION_COUNT_EXT"; + case -1059: return "CL_INVALID_PARTITION_NAME_EXT"; + case -1094: return "CL_INVALID_ACCELERATOR_INTEL"; + case -1095: return "CL_INVALID_ACCELERATOR_TYPE_INTEL"; + case -1096: return "CL_INVALID_ACCELERATOR_DESCRIPTOR_INTEL"; + case -1097: return "CL_ACCELERATOR_TYPE_NOT_SUPPORTED_INTEL"; + case -1000: return "CL_INVALID_GL_SHAREGROUP_REFERENCE_KHR"; + case -1098: return "CL_INVALID_VA_API_MEDIA_ADAPTER_INTEL"; + case -1099: return "CL_INVALID_VA_API_MEDIA_SURFACE_INTEL"; + case -1100: return "CL_VA_API_MEDIA_SURFACE_ALREADY_ACQUIRED_INTEL"; + case -1101: return "CL_VA_API_MEDIA_SURFACE_NOT_ACQUIRED_INTEL"; + default: return "CL_UNKNOWN_ERROR"; + } +} + +OpenClWrapper::OpenClWrapper() +{ + m_camera_id = 0; //cmd.get("camera"); + m_file_name = "0"; //cmd.get("video"); + + m_is_initialized = false; + m_running = false; + m_process = false; + m_use_buffer = false; + + m_t0 = 0; + m_t1 = 0; + m_time = 0.0; + m_frequency = (float)cv::getTickFrequency(); + + m_context = 0; + //m_device_id = 0; + m_queue = 0; + m_program = 0; + m_kernelBuf = 0; + m_kernelImg = 0; + m_img_src = 0; + m_mem_obj = 0; +} // ctor + + +OpenClWrapper::~OpenClWrapper() +{ + if (m_queue) + { + clFinish(m_queue); + clReleaseCommandQueue(m_queue); + m_queue = 0; + } + + if (m_program) + { + clReleaseProgram(m_program); + m_program = 0; + } + + if (m_img_src) + { + clReleaseMemObject(m_img_src); + m_img_src = 0; + } + + if (m_mem_obj) + { + clReleaseMemObject(m_mem_obj); + m_mem_obj = 0; + } + + if (m_kernelBuf) + { + clReleaseKernel(m_kernelBuf); + m_kernelBuf = 0; + } + + if (m_kernelImg) + { + clReleaseKernel(m_kernelImg); + m_kernelImg = 0; + } + + /*if (m_device_id) + { + clReleaseDevice(m_device_id); + m_device_id = 0; + }*/ + + if (m_context) + { + clReleaseContext(m_context); + m_context = 0; + } +} // dtor + +bool OpenClWrapper::m_is_initialized = false; +cl_context OpenClWrapper::m_context = nullptr; +cl_command_queue OpenClWrapper::m_queue = nullptr; + +int OpenClWrapper::initOpenCL() +{ + if (m_is_initialized){ + std::cout << "OpenCL already initialized." << std::endl; + return CL_SUCCESS; + } + + cl_int res = CL_SUCCESS; + cl_uint num_entries = 0; + + res = clGetPlatformIDs(0, 0, &num_entries); + if (CL_SUCCESS != res) + return -1; + + opencl::PlatformInfo m_platformInfo; + opencl::DeviceInfo m_deviceInfo; + cl_device_id m_device_id; + + std::vector m_platform_ids; + + m_platform_ids.resize(num_entries); + + res = clGetPlatformIDs(num_entries, &m_platform_ids[0], 0); + if (CL_SUCCESS != res) + return -1; + + unsigned int i; + + // create context from first platform with GPU device + for (i = 0; i < m_platform_ids.size(); i++) + { + cl_context_properties props[] = + { + CL_CONTEXT_PLATFORM, + (cl_context_properties)(m_platform_ids[i]), + 0 + }; + + cl_device_id device; + // Get the first available device on the platform + res = clGetDeviceIDs(m_platform_ids[i], CL_DEVICE_TYPE_GPU, 1, &device, NULL); + if (res != CL_SUCCESS) { + printf("Error getting device IDs\n"); + return 1; + } + + // Query the device for supported image formats + cl_image_format supportedFormats[128]; // Assuming maximum 128 supported formats + cl_uint numSupportedFormats; + + m_context = clCreateContextFromType(props, CL_DEVICE_TYPE_GPU, 0, 0, &res); + if (0 == m_context || CL_SUCCESS != res) + continue; + + res = clGetContextInfo(m_context, CL_CONTEXT_DEVICES, sizeof(cl_device_id), &m_device_id, 0); + if (CL_SUCCESS != res) + return -1; + + res = clGetSupportedImageFormats(m_context, CL_MEM_READ_WRITE, CL_MEM_OBJECT_IMAGE2D, 128, supportedFormats, &numSupportedFormats); + if (res != CL_SUCCESS) { + printf("Error getting supported image formats\n"); + return 1; + } + + // Check if CL_HALF_FLOAT is supported + cl_bool supportsHalfFloat = CL_FALSE; + for (cl_uint i = 0; i < numSupportedFormats; ++i) { + if (supportedFormats[i].image_channel_data_type == CL_HALF_FLOAT) { + supportsHalfFloat = CL_TRUE; + break; + } + } + + // Print the result + if (supportsHalfFloat == CL_TRUE) { + printf("Device supports CL_HALF_FLOAT format\n"); + } else { + printf("Device does not support CL_HALF_FLOAT format\n"); + } + + m_queue = clCreateCommandQueue(m_context, m_device_id, 0, &res); + if (0 == m_queue || CL_SUCCESS != res) + return -1; + + m_platformInfo.QueryInfo(m_platform_ids[i]); + m_deviceInfo.QueryInfo(m_device_id); + + // attach OpenCL context to OpenCV + cv::ocl::attachContext(m_platformInfo.Name(), m_platform_ids[i], m_context, m_device_id); + + break; + } + + m_is_initialized = true; + std::cout << "OpenCL initialized for the first time." << std::endl; + cout << "Version : " << m_platformInfo.Version() << std::endl; + cout << "Name : " << m_platformInfo.Name()<< std::endl; + cout << "Device : " << m_deviceInfo.Name()<< std::endl; + + if (m_device_id) + { + clReleaseDevice(m_device_id); + m_device_id = 0; + } + + //printInfo(); + return m_context != 0 ? CL_SUCCESS : -1; +} // initOpenCL() + +int OpenClWrapper::createMemObject(cl_mem* mem_obj, cv::UMat& inputData){ + //cv::ocl::Image2D image = cv::ocl::Image2D(inputData); + //mem_obj[0] = image.handle; + + return 0; +} +int OpenClWrapper::createMemObject(cl_mem* mem_obj, cv::Mat& inputData){ + // OLD IMPLEMENTATION + cl_int res = CL_SUCCESS; + cl_mem mem = mem_obj[0]; + + if (inputData.ptr() == nullptr) + { + std::cout << "Error:createMemObject nupptr as input ptr" << std::endl; + return -1; + } + + cl_image_format fmt; + cl_channel_order channelOrder = GetChannelOrderFromMatType(inputData.type()); + fmt.image_channel_order = channelOrder; + cl_channel_type channelType = GetChannelDataTypeFromOrder(channelOrder); + fmt.image_channel_data_type = channelType; + + cl_mem_flags flags_dst = CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR; + // TODO: fix CL_INVALID_IMAGE_FORMAT_DESCRIPTOR now - Hardcoded values + // fmt.image_channel_order = CL_R; + // fmt.image_channel_data_type = CL_UNSIGNED_INT8; + + cl_image_desc desc_dst; + memset(&desc_dst, 0, sizeof(cl_image_desc)); + desc_dst.image_type = CL_MEM_OBJECT_IMAGE2D; + desc_dst.image_width = inputData.cols; + desc_dst.image_height = inputData.rows; + desc_dst.image_depth = 0; + desc_dst.image_array_size = 0; + // desc_dst.image_row_pitch = inputData.step[0]; + desc_dst.image_row_pitch = inputData.step[0]; + desc_dst.image_slice_pitch = 0; + desc_dst.num_mip_levels = 0; + desc_dst.num_samples = 0; + desc_dst.buffer = 0; + mem = clCreateImage(m_context, flags_dst, &fmt, &desc_dst, inputData.ptr(), &res); + if (0 == mem || CL_SUCCESS != res){ + std::cout <<"Error: " << clGetErrorString(res) << std::endl; + return -1; + } + + mem_obj[0] = mem; + + return 0; +} + +// this function is an example of "typical" OpenCL processing pipeline +// It creates OpenCL buffer or image, depending on use_buffer flag, +// from input media frame and process these data +// (inverts each pixel value in half of frame) with OpenCL kernel +int OpenClWrapper::process_frame_with_open_cl(cv::Mat& frame, bool use_buffer, cl_mem* mem_obj) +{ + cl_int res = CL_SUCCESS; + + CV_Assert(mem_obj); + + cl_kernel kernel = 0; + cl_mem mem = mem_obj[0]; + + if (0 == mem || 0 == m_img_src) + { + // allocate/delete cl memory objects every frame for the simplicity. + // in real application more efficient pipeline can be built. + + if (use_buffer) + { + cl_mem_flags flags = CL_MEM_READ_WRITE | CL_MEM_USE_HOST_PTR; + + mem = clCreateBuffer(m_context, flags, frame.total(), frame.ptr(), &res); + if (0 == mem || CL_SUCCESS != res) + return -1; + + res = clSetKernelArg(m_kernelBuf, 0, sizeof(cl_mem), &mem); + if (CL_SUCCESS != res) + return -1; + + res = clSetKernelArg(m_kernelBuf, 1, sizeof(int), &frame.step[0]); + if (CL_SUCCESS != res) + return -1; + + res = clSetKernelArg(m_kernelBuf, 2, sizeof(int), &frame.rows); + if (CL_SUCCESS != res) + return -1; + + int cols2 = frame.cols / 2; + res = clSetKernelArg(m_kernelBuf, 3, sizeof(int), &cols2); + if (CL_SUCCESS != res) + return -1; + + kernel = m_kernelBuf; + } + else + { + cl_mem_flags flags_src = CL_MEM_READ_ONLY | CL_MEM_USE_HOST_PTR; + + cl_image_format fmt; + fmt.image_channel_order = CL_R; + fmt.image_channel_data_type = CL_UNSIGNED_INT8; + + cl_image_desc desc_src; + desc_src.image_type = CL_MEM_OBJECT_IMAGE2D; + desc_src.image_width = frame.cols; + desc_src.image_height = frame.rows; + desc_src.image_depth = 0; + desc_src.image_array_size = 0; + desc_src.image_row_pitch = frame.step[0]; + desc_src.image_slice_pitch = 0; + desc_src.num_mip_levels = 0; + desc_src.num_samples = 0; + desc_src.buffer = 0; + m_img_src = clCreateImage(m_context, flags_src, &fmt, &desc_src, frame.ptr(), &res); + if (0 == m_img_src || CL_SUCCESS != res) + return -1; + + cl_mem_flags flags_dst = CL_MEM_READ_WRITE | CL_MEM_ALLOC_HOST_PTR; + + cl_image_desc desc_dst; + desc_dst.image_type = CL_MEM_OBJECT_IMAGE2D; + desc_dst.image_width = frame.cols; + desc_dst.image_height = frame.rows; + desc_dst.image_depth = 0; + desc_dst.image_array_size = 0; + desc_dst.image_row_pitch = 0; + desc_dst.image_slice_pitch = 0; + desc_dst.num_mip_levels = 0; + desc_dst.num_samples = 0; + desc_dst.buffer = 0; + mem = clCreateImage(m_context, flags_dst, &fmt, &desc_dst, 0, &res); + if (0 == mem || CL_SUCCESS != res) + return -1; + + size_t origin[] = { 0, 0, 0 }; + size_t region[] = { (size_t)frame.cols, (size_t)frame.rows, 1 }; + cl_event asyncEvent = 0; + res = clEnqueueCopyImage(m_queue, m_img_src, mem, origin, origin, region, 0, 0, &asyncEvent); + if (CL_SUCCESS != res) + return -1; + + res = clWaitForEvents(1, &asyncEvent); + clReleaseEvent(asyncEvent); + if (CL_SUCCESS != res) + return -1; + + res = clSetKernelArg(m_kernelImg, 0, sizeof(cl_mem), &m_img_src); + if (CL_SUCCESS != res) + return -1; + + res = clSetKernelArg(m_kernelImg, 1, sizeof(cl_mem), &mem); + if (CL_SUCCESS != res) + return -1; + + kernel = m_kernelImg; + } + } + + // process left half of frame in OpenCL + size_t size[] = { (size_t)frame.cols / 2, (size_t)frame.rows }; + cl_event asyncEvent = 0; + res = clEnqueueNDRangeKernel(m_queue, kernel, 2, 0, size, 0, 0, 0, &asyncEvent); + if (CL_SUCCESS != res) + return -1; + + res = clWaitForEvents(1, &asyncEvent); + clReleaseEvent(asyncEvent); + if (CL_SUCCESS != res) + return -1; + + mem_obj[0] = mem; + + return 0; +} + + +// this function is an example of interoperability between OpenCL buffer +// and OpenCV UMat objects. It converts (without copying data) OpenCL buffer +// to OpenCV UMat and then do blur on these data +int OpenClWrapper::process_cl_buffer_with_opencv(cl_mem buffer, size_t step, int rows, int cols, int type, cv::UMat& u) +{ + cv::ocl::convertFromBuffer(buffer, step, rows, cols, type, u); + + // process right half of frame in OpenCV + cv::Point pt(u.cols / 2, 0); + cv::Size sz(u.cols / 2, u.rows); + cv::Rect roi(pt, sz); + cv::UMat uroi(u, roi); + cv::blur(uroi, uroi, cv::Size(7, 7), cv::Point(-3, -3)); + + if (buffer) + clReleaseMemObject(buffer); + m_mem_obj = 0; + + return 0; +} + + +// this function is an example of interoperability between OpenCL image +// and OpenCV UMat objects. It converts OpenCL image +// to OpenCV UMat and then do blur on these data +int OpenClWrapper::process_cl_image_with_opencv(cl_mem image, cv::UMat& u) +{ + cv::ocl::convertFromImage(image, u); + + // process right half of frame in OpenCV + cv::Point pt(u.cols / 2, 0); + cv::Size sz(u.cols / 2, u.rows); + cv::Rect roi(pt, sz); + cv::UMat uroi(u, roi); + cv::blur(uroi, uroi, cv::Size(7, 7), cv::Point(-3, -3)); + + if (image) + clReleaseMemObject(image); + m_mem_obj = 0; + + if (m_img_src) + clReleaseMemObject(m_img_src); + m_img_src = 0; + + return 0; +} + + +int OpenClWrapper::run() +{ + if (0 != initOpenCL()) + return -1; + + //if (0 != initVideoSource()) + // return -1; + + Mat img_to_show; + + // set running state until ESC pressed + setRunning(true); + // set process flag to show some data processing + // can be toggled on/off by 'p' button + setDoProcess(true); + // set use buffer flag, + // when it is set to true, will demo interop opencl buffer and cv::Umat, + // otherwise demo interop opencl image and cv::UMat + // can be switched on/of by SPACE button + setUseBuffer(true); + + // Iterate over all frames + while (isRunning() && nextFrame(m_frame)) + { + cv::cvtColor(m_frame, m_frameGray, COLOR_BGR2GRAY); + + UMat uframe; + + // work + //timerStart(); + + if (doProcess()) + { + process_frame_with_open_cl(m_frameGray, useBuffer(), &m_mem_obj); + + if (useBuffer()) + process_cl_buffer_with_opencv( + m_mem_obj, m_frameGray.step[0], m_frameGray.rows, m_frameGray.cols, m_frameGray.type(), uframe); + else + process_cl_image_with_opencv(m_mem_obj, uframe); + } + else + { + m_frameGray.copyTo(uframe); + } + + //timerEnd(); + + uframe.copyTo(img_to_show); + + imshow("opencl_interop", img_to_show); + + //handleKey((char)waitKey(3)); + } + + return 0; +} + +void OpenClWrapper::printInfo() +{ + // TODO +} \ No newline at end of file diff --git a/mediapipe/framework/formats/helpers.hpp b/mediapipe/framework/formats/helpers.hpp new file mode 100644 index 0000000000..3ab16cfd43 --- /dev/null +++ b/mediapipe/framework/formats/helpers.hpp @@ -0,0 +1,469 @@ +/* +// The example of interoperability between OpenCL and OpenCV. +// This will loop through frames of video either from input media file +// or camera device and do processing of these data in OpenCL and then +// in OpenCV. In OpenCL it does inversion of pixels in left half of frame and +// in OpenCV it does blurring in the right half of frame. +*/ +#include +#include +#include +#include +#include +#include +#include +#include + +#define CL_USE_DEPRECATED_OPENCL_1_1_APIS +#define CL_USE_DEPRECATED_OPENCL_1_2_APIS +#define CL_USE_DEPRECATED_OPENCL_2_0_APIS // eliminate build warning +#define CL_TARGET_OPENCL_VERSION 200 // 2.0 + +#ifdef __APPLE__ +#define CL_SILENCE_DEPRECATION +#include +#else +#include +#endif + +#include +#include +#include +#include +#include + + +using namespace std; +using namespace cv; + +namespace opencl { + +class PlatformInfo +{ +public: + PlatformInfo() + {} + + ~PlatformInfo() + {} + + cl_int QueryInfo(cl_platform_id id) + { + query_param(id, CL_PLATFORM_PROFILE, m_profile); + query_param(id, CL_PLATFORM_VERSION, m_version); + query_param(id, CL_PLATFORM_NAME, m_name); + query_param(id, CL_PLATFORM_VENDOR, m_vendor); + query_param(id, CL_PLATFORM_EXTENSIONS, m_extensions); + return CL_SUCCESS; + } + + std::string Profile() { return m_profile; } + std::string Version() { return m_version; } + std::string Name() { return m_name; } + std::string Vendor() { return m_vendor; } + std::string Extensions() { return m_extensions; } + +private: + cl_int query_param(cl_platform_id id, cl_platform_info param, std::string& paramStr) + { + cl_int res; + + size_t psize; + cv::AutoBuffer buf; + + res = clGetPlatformInfo(id, param, 0, 0, &psize); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetPlatformInfo failed")); + + buf.resize(psize); + res = clGetPlatformInfo(id, param, psize, buf, 0); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetPlatformInfo failed")); + + // just in case, ensure trailing zero for ASCIIZ string + buf[psize] = 0; + + paramStr = buf; + + return CL_SUCCESS; + } + +private: + std::string m_profile; + std::string m_version; + std::string m_name; + std::string m_vendor; + std::string m_extensions; +}; + + +class DeviceInfo +{ +public: + DeviceInfo() + {} + + ~DeviceInfo() + {} + + cl_int QueryInfo(cl_device_id id) + { + query_param(id, CL_DEVICE_TYPE, m_type); + query_param(id, CL_DEVICE_VENDOR_ID, m_vendor_id); + query_param(id, CL_DEVICE_MAX_COMPUTE_UNITS, m_max_compute_units); + query_param(id, CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS, m_max_work_item_dimensions); + query_param(id, CL_DEVICE_MAX_WORK_ITEM_SIZES, m_max_work_item_sizes); + query_param(id, CL_DEVICE_MAX_WORK_GROUP_SIZE, m_max_work_group_size); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR, m_preferred_vector_width_char); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_SHORT, m_preferred_vector_width_short); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_INT, m_preferred_vector_width_int); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_LONG, m_preferred_vector_width_long); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT, m_preferred_vector_width_float); + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE, m_preferred_vector_width_double); +#if defined(CL_VERSION_1_1) + query_param(id, CL_DEVICE_PREFERRED_VECTOR_WIDTH_HALF, m_preferred_vector_width_half); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_CHAR, m_native_vector_width_char); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_SHORT, m_native_vector_width_short); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_INT, m_native_vector_width_int); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_LONG, m_native_vector_width_long); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_FLOAT, m_native_vector_width_float); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE, m_native_vector_width_double); + query_param(id, CL_DEVICE_NATIVE_VECTOR_WIDTH_HALF, m_native_vector_width_half); +#endif + query_param(id, CL_DEVICE_MAX_CLOCK_FREQUENCY, m_max_clock_frequency); + query_param(id, CL_DEVICE_ADDRESS_BITS, m_address_bits); + query_param(id, CL_DEVICE_MAX_MEM_ALLOC_SIZE, m_max_mem_alloc_size); + query_param(id, CL_DEVICE_IMAGE_SUPPORT, m_image_support); + query_param(id, CL_DEVICE_MAX_READ_IMAGE_ARGS, m_max_read_image_args); + query_param(id, CL_DEVICE_MAX_WRITE_IMAGE_ARGS, m_max_write_image_args); +#if defined(CL_VERSION_2_0) + query_param(id, CL_DEVICE_MAX_READ_WRITE_IMAGE_ARGS, m_max_read_write_image_args); +#endif + query_param(id, CL_DEVICE_IMAGE2D_MAX_WIDTH, m_image2d_max_width); + query_param(id, CL_DEVICE_IMAGE2D_MAX_HEIGHT, m_image2d_max_height); + query_param(id, CL_DEVICE_IMAGE3D_MAX_WIDTH, m_image3d_max_width); + query_param(id, CL_DEVICE_IMAGE3D_MAX_HEIGHT, m_image3d_max_height); + query_param(id, CL_DEVICE_IMAGE3D_MAX_DEPTH, m_image3d_max_depth); +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_IMAGE_MAX_BUFFER_SIZE, m_image_max_buffer_size); + query_param(id, CL_DEVICE_IMAGE_MAX_ARRAY_SIZE, m_image_max_array_size); +#endif + query_param(id, CL_DEVICE_MAX_SAMPLERS, m_max_samplers); +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_IMAGE_PITCH_ALIGNMENT, m_image_pitch_alignment); + query_param(id, CL_DEVICE_IMAGE_BASE_ADDRESS_ALIGNMENT, m_image_base_address_alignment); +#endif +#if defined(CL_VERSION_2_0) + query_param(id, CL_DEVICE_MAX_PIPE_ARGS, m_max_pipe_args); + query_param(id, CL_DEVICE_PIPE_MAX_ACTIVE_RESERVATIONS, m_pipe_max_active_reservations); + query_param(id, CL_DEVICE_PIPE_MAX_PACKET_SIZE, m_pipe_max_packet_size); +#endif + query_param(id, CL_DEVICE_MAX_PARAMETER_SIZE, m_max_parameter_size); + query_param(id, CL_DEVICE_MEM_BASE_ADDR_ALIGN, m_mem_base_addr_align); + query_param(id, CL_DEVICE_SINGLE_FP_CONFIG, m_single_fp_config); +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_DOUBLE_FP_CONFIG, m_double_fp_config); +#endif + query_param(id, CL_DEVICE_GLOBAL_MEM_CACHE_TYPE, m_global_mem_cache_type); + query_param(id, CL_DEVICE_GLOBAL_MEM_CACHELINE_SIZE, m_global_mem_cacheline_size); + query_param(id, CL_DEVICE_GLOBAL_MEM_CACHE_SIZE, m_global_mem_cache_size); + query_param(id, CL_DEVICE_GLOBAL_MEM_SIZE, m_global_mem_size); + query_param(id, CL_DEVICE_MAX_CONSTANT_BUFFER_SIZE, m_max_constant_buffer_size); + query_param(id, CL_DEVICE_MAX_CONSTANT_ARGS, m_max_constant_args); +#if defined(CL_VERSION_2_0) + query_param(id, CL_DEVICE_MAX_GLOBAL_VARIABLE_SIZE, m_max_global_variable_size); + query_param(id, CL_DEVICE_GLOBAL_VARIABLE_PREFERRED_TOTAL_SIZE, m_global_variable_preferred_total_size); +#endif + query_param(id, CL_DEVICE_LOCAL_MEM_TYPE, m_local_mem_type); + query_param(id, CL_DEVICE_LOCAL_MEM_SIZE, m_local_mem_size); + query_param(id, CL_DEVICE_ERROR_CORRECTION_SUPPORT, m_error_correction_support); +#if defined(CL_VERSION_1_1) + query_param(id, CL_DEVICE_HOST_UNIFIED_MEMORY, m_host_unified_memory); +#endif + query_param(id, CL_DEVICE_PROFILING_TIMER_RESOLUTION, m_profiling_timer_resolution); + query_param(id, CL_DEVICE_ENDIAN_LITTLE, m_endian_little); + query_param(id, CL_DEVICE_AVAILABLE, m_available); + query_param(id, CL_DEVICE_COMPILER_AVAILABLE, m_compiler_available); +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_LINKER_AVAILABLE, m_linker_available); +#endif + query_param(id, CL_DEVICE_EXECUTION_CAPABILITIES, m_execution_capabilities); + query_param(id, CL_DEVICE_QUEUE_PROPERTIES, m_queue_properties); +#if defined(CL_VERSION_2_0) + query_param(id, CL_DEVICE_QUEUE_ON_HOST_PROPERTIES, m_queue_on_host_properties); + query_param(id, CL_DEVICE_QUEUE_ON_DEVICE_PROPERTIES, m_queue_on_device_properties); + query_param(id, CL_DEVICE_QUEUE_ON_DEVICE_PREFERRED_SIZE, m_queue_on_device_preferred_size); + query_param(id, CL_DEVICE_QUEUE_ON_DEVICE_MAX_SIZE, m_queue_on_device_max_size); + query_param(id, CL_DEVICE_MAX_ON_DEVICE_QUEUES, m_max_on_device_queues); + query_param(id, CL_DEVICE_MAX_ON_DEVICE_EVENTS, m_max_on_device_events); +#endif +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_BUILT_IN_KERNELS, m_built_in_kernels); +#endif + query_param(id, CL_DEVICE_PLATFORM, m_platform); + query_param(id, CL_DEVICE_NAME, m_name); + query_param(id, CL_DEVICE_VENDOR, m_vendor); + query_param(id, CL_DRIVER_VERSION, m_driver_version); + query_param(id, CL_DEVICE_PROFILE, m_profile); + query_param(id, CL_DEVICE_VERSION, m_version); +#if defined(CL_VERSION_1_1) + query_param(id, CL_DEVICE_OPENCL_C_VERSION, m_opencl_c_version); +#endif + query_param(id, CL_DEVICE_EXTENSIONS, m_extensions); +#if defined(CL_VERSION_1_2) + query_param(id, CL_DEVICE_PRINTF_BUFFER_SIZE, m_printf_buffer_size); + query_param(id, CL_DEVICE_PREFERRED_INTEROP_USER_SYNC, m_preferred_interop_user_sync); + query_param(id, CL_DEVICE_PARENT_DEVICE, m_parent_device); + query_param(id, CL_DEVICE_PARTITION_MAX_SUB_DEVICES, m_partition_max_sub_devices); + query_param(id, CL_DEVICE_PARTITION_PROPERTIES, m_partition_properties); + query_param(id, CL_DEVICE_PARTITION_AFFINITY_DOMAIN, m_partition_affinity_domain); + query_param(id, CL_DEVICE_PARTITION_TYPE, m_partition_type); + query_param(id, CL_DEVICE_REFERENCE_COUNT, m_reference_count); +#endif + return CL_SUCCESS; + } + + std::string Name() { return m_name; } + +private: + template + cl_int query_param(cl_device_id id, cl_device_info param, T& value) + { + cl_int res; + size_t size = 0; + + res = clGetDeviceInfo(id, param, 0, 0, &size); + if (CL_SUCCESS != res && size != 0) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + if (0 == size) + return CL_SUCCESS; + + if (sizeof(T) != size) + throw std::runtime_error(std::string("clGetDeviceInfo: param size mismatch")); + + res = clGetDeviceInfo(id, param, size, &value, 0); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + return CL_SUCCESS; + } + + template + cl_int query_param(cl_device_id id, cl_device_info param, std::vector& value) + { + cl_int res; + size_t size; + + res = clGetDeviceInfo(id, param, 0, 0, &size); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + if (0 == size) + return CL_SUCCESS; + + value.resize(size / sizeof(T)); + + res = clGetDeviceInfo(id, param, size, &value[0], 0); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + return CL_SUCCESS; + } + + cl_int query_param(cl_device_id id, cl_device_info param, std::string& value) + { + cl_int res; + size_t size; + + res = clGetDeviceInfo(id, param, 0, 0, &size); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + value.resize(size + 1); + + res = clGetDeviceInfo(id, param, size, &value[0], 0); + if (CL_SUCCESS != res) + throw std::runtime_error(std::string("clGetDeviceInfo failed")); + + // just in case, ensure trailing zero for ASCIIZ string + value[size] = 0; + + return CL_SUCCESS; + } + +private: + cl_device_type m_type; + cl_uint m_vendor_id; + cl_uint m_max_compute_units; + cl_uint m_max_work_item_dimensions; + std::vector m_max_work_item_sizes; + size_t m_max_work_group_size; + cl_uint m_preferred_vector_width_char; + cl_uint m_preferred_vector_width_short; + cl_uint m_preferred_vector_width_int; + cl_uint m_preferred_vector_width_long; + cl_uint m_preferred_vector_width_float; + cl_uint m_preferred_vector_width_double; +#if defined(CL_VERSION_1_1) + cl_uint m_preferred_vector_width_half; + cl_uint m_native_vector_width_char; + cl_uint m_native_vector_width_short; + cl_uint m_native_vector_width_int; + cl_uint m_native_vector_width_long; + cl_uint m_native_vector_width_float; + cl_uint m_native_vector_width_double; + cl_uint m_native_vector_width_half; +#endif + cl_uint m_max_clock_frequency; + cl_uint m_address_bits; + cl_ulong m_max_mem_alloc_size; + cl_bool m_image_support; + cl_uint m_max_read_image_args; + cl_uint m_max_write_image_args; +#if defined(CL_VERSION_2_0) + cl_uint m_max_read_write_image_args; +#endif + size_t m_image2d_max_width; + size_t m_image2d_max_height; + size_t m_image3d_max_width; + size_t m_image3d_max_height; + size_t m_image3d_max_depth; +#if defined(CL_VERSION_1_2) + size_t m_image_max_buffer_size; + size_t m_image_max_array_size; +#endif + cl_uint m_max_samplers; +#if defined(CL_VERSION_1_2) + cl_uint m_image_pitch_alignment; + cl_uint m_image_base_address_alignment; +#endif +#if defined(CL_VERSION_2_0) + cl_uint m_max_pipe_args; + cl_uint m_pipe_max_active_reservations; + cl_uint m_pipe_max_packet_size; +#endif + size_t m_max_parameter_size; + cl_uint m_mem_base_addr_align; + cl_device_fp_config m_single_fp_config; +#if defined(CL_VERSION_1_2) + cl_device_fp_config m_double_fp_config; +#endif + cl_device_mem_cache_type m_global_mem_cache_type; + cl_uint m_global_mem_cacheline_size; + cl_ulong m_global_mem_cache_size; + cl_ulong m_global_mem_size; + cl_ulong m_max_constant_buffer_size; + cl_uint m_max_constant_args; +#if defined(CL_VERSION_2_0) + size_t m_max_global_variable_size; + size_t m_global_variable_preferred_total_size; +#endif + cl_device_local_mem_type m_local_mem_type; + cl_ulong m_local_mem_size; + cl_bool m_error_correction_support; +#if defined(CL_VERSION_1_1) + cl_bool m_host_unified_memory; +#endif + size_t m_profiling_timer_resolution; + cl_bool m_endian_little; + cl_bool m_available; + cl_bool m_compiler_available; +#if defined(CL_VERSION_1_2) + cl_bool m_linker_available; +#endif + cl_device_exec_capabilities m_execution_capabilities; + cl_command_queue_properties m_queue_properties; +#if defined(CL_VERSION_2_0) + cl_command_queue_properties m_queue_on_host_properties; + cl_command_queue_properties m_queue_on_device_properties; + cl_uint m_queue_on_device_preferred_size; + cl_uint m_queue_on_device_max_size; + cl_uint m_max_on_device_queues; + cl_uint m_max_on_device_events; +#endif +#if defined(CL_VERSION_1_2) + std::string m_built_in_kernels; +#endif + cl_platform_id m_platform; + std::string m_name; + std::string m_vendor; + std::string m_driver_version; + std::string m_profile; + std::string m_version; +#if defined(CL_VERSION_1_1) + std::string m_opencl_c_version; +#endif + std::string m_extensions; +#if defined(CL_VERSION_1_2) + size_t m_printf_buffer_size; + cl_bool m_preferred_interop_user_sync; + cl_device_id m_parent_device; + cl_uint m_partition_max_sub_devices; + std::vector m_partition_properties; + cl_device_affinity_domain m_partition_affinity_domain; + std::vector m_partition_type; + cl_uint m_reference_count; +#endif +}; + +} // namespace opencl + +void dumpMatToFile(std::string& fileName, cv::UMat& umat); +void dumpMatToFile(std::string& fileName, cv::Mat& umat); + +class OpenClWrapper +{ +public: + OpenClWrapper(); + ~OpenClWrapper(); + + static int initOpenCL(); + + int createMemObject(cl_mem* mem_obj, cv::UMat& inputData); + int createMemObject(cl_mem* mem_obj, cv::Mat& inputData); + int process_frame_with_open_cl(cv::Mat& frame, bool use_buffer, cl_mem* cl_buffer); + int process_cl_buffer_with_opencv(cl_mem buffer, size_t step, int rows, int cols, int type, cv::UMat& u); + int process_cl_image_with_opencv(cl_mem image, cv::UMat& u); + + int run(); + + bool isRunning() { return m_running; } + bool doProcess() { return m_process; } + bool useBuffer() { return m_use_buffer; } + + void setRunning(bool running) { m_running = running; } + void setDoProcess(bool process) { m_process = process; } + void setUseBuffer(bool use_buffer) { m_use_buffer = use_buffer; } + void printInfo(); + + cl_mem m_mem_obj; +protected: + bool nextFrame(cv::Mat& frame) { return m_cap.read(frame); } + std::string message() const; + +private: + bool m_running; + bool m_process; + bool m_use_buffer; + + int64 m_t0; + int64 m_t1; + float m_time; + float m_frequency; + + string m_file_name; + int m_camera_id; + cv::VideoCapture m_cap; + cv::Mat m_frame; + cv::Mat m_frameGray; + + static bool m_is_initialized; + static cl_context m_context; + static cl_command_queue m_queue; + + cl_program m_program; + cl_kernel m_kernelBuf; + cl_kernel m_kernelImg; + cl_mem m_img_src; // used as src in case processing of cl image +}; \ No newline at end of file diff --git a/mediapipe/framework/formats/image_frame.cc b/mediapipe/framework/formats/image_frame.cc index 2de819a35b..61ba25a86e 100644 --- a/mediapipe/framework/formats/image_frame.cc +++ b/mediapipe/framework/formats/image_frame.cc @@ -20,6 +20,8 @@ #include #include +#include +#include #include #include @@ -58,22 +60,40 @@ const uint32_t ImageFrame::kDefaultAlignmentBoundary; const uint32_t ImageFrame::kGlDefaultAlignmentBoundary; ImageFrame::ImageFrame() - : format_(ImageFormat::UNKNOWN), width_(0), height_(0), width_step_(0) {} + : format_(ImageFormat::UNKNOWN), width_(0), height_(0), width_step_(0) { + //CHECK_NE(ocl.initOpenCL(), -1); + } + +ImageFrame::ImageFrame(cv::UMat& inputData, ImageFormat::Format format, int width, int height, + uint32_t alignment_boundary) + : format_(format), width_(width), height_(height) { + //CHECK_NE(ocl.initOpenCL(), -1); + Reset(inputData, format, width, height, alignment_boundary); +} ImageFrame::ImageFrame(ImageFormat::Format format, int width, int height, uint32_t alignment_boundary) : format_(format), width_(width), height_(height) { + //CHECK_NE(ocl.initOpenCL(), -1); Reset(format, width, height, alignment_boundary); } ImageFrame::ImageFrame(ImageFormat::Format format, int width, int height) : format_(format), width_(width), height_(height) { + //CHECK_NE(ocl.initOpenCL(), -1); Reset(format, width, height, kDefaultAlignmentBoundary); } +ImageFrame::ImageFrame(cv::UMat& inputData, ImageFormat::Format format, int width, int height) + : format_(format), width_(width), height_(height) { + //CHECK_NE(ocl.initOpenCL(), -1); + Reset(inputData, format, width, height, kDefaultAlignmentBoundary); +} + ImageFrame::ImageFrame(ImageFormat::Format format, int width, int height, int width_step, uint8_t* pixel_data, ImageFrame::Deleter deleter) { + //CHECK_NE(ocl.initOpenCL(), -1); AdoptPixelData(format, width, height, width_step, pixel_data, deleter); } @@ -85,6 +105,7 @@ ImageFrame& ImageFrame::operator=(ImageFrame&& move_from) { width_ = move_from.width_; height_ = move_from.height_; width_step_ = move_from.width_step_; + //ocl = move_from.ocl; move_from.format_ = ImageFormat::UNKNOWN; move_from.width_ = 0; @@ -93,11 +114,28 @@ ImageFrame& ImageFrame::operator=(ImageFrame&& move_from) { return *this; } +void ImageFrame::Reset(cv::UMat& inputData, ImageFormat::Format format, int width, int height, + uint32_t alignment_boundary) { + format_ = format; + width_ = width; + height_ = height; + CHECK_NE(ImageFormat::UNKNOWN, format_); + CHECK(IsValidAlignmentNumber(alignment_boundary)); + width_step_ = width * NumberOfChannels() * ByteDepth(); + + width_step_ = ((width_step_ - 1) | (alignment_boundary - 1)) + 1; + //CHECK_NE(ocl.createMemObject(&ocl.m_mem_obj, inputData), -1); + + pixel_data_ = { reinterpret_cast(inputData.handle(ACCESS_RW)) , PixelDataDeleter::kNone}; + } + void ImageFrame::Reset(ImageFormat::Format format, int width, int height, uint32_t alignment_boundary) { format_ = format; width_ = width; height_ = height; + // TODO graph execution check + CHECK_NE(0,0); CHECK_NE(ImageFormat::UNKNOWN, format_); CHECK(IsValidAlignmentNumber(alignment_boundary)); width_step_ = width * NumberOfChannels() * ByteDepth(); diff --git a/mediapipe/framework/formats/image_frame.h b/mediapipe/framework/formats/image_frame.h index 6fcefbd38f..ce7ea974f5 100644 --- a/mediapipe/framework/formats/image_frame.h +++ b/mediapipe/framework/formats/image_frame.h @@ -43,6 +43,7 @@ #include "mediapipe/framework/port.h" #include "mediapipe/framework/port/integral_types.h" #include "mediapipe/framework/tool/type_util.h" +#include "mediapipe/framework/formats/helpers.hpp" #define IMAGE_FRAME_RAW_IMAGE MEDIAPIPE_HAS_RTTI @@ -61,6 +62,7 @@ namespace mediapipe { // // Do not assume that the pixel data is stored contiguously. It may be // stored with row padding for alignment purposes. + class ImageFrame { public: typedef std::function Deleter; @@ -105,8 +107,11 @@ class ImageFrame { // be stored contiguously). ImageFrame(ImageFormat::Format format, int width, int height, uint32 alignment_boundary); + ImageFrame(cv::UMat& inputData, ImageFormat::Format format, int width, int height, + uint32_t alignment_boundary); // Same as above, but use kDefaultAlignmentBoundary for alignment_boundary. ImageFrame(ImageFormat::Format format, int width, int height); + ImageFrame(cv::UMat& inputData, ImageFormat::Format format, int width, int height); // Acquires ownership of pixel_data. Sets the deletion method // to use on pixel_data with deletion_method (which defaults @@ -209,6 +214,9 @@ class ImageFrame { void Reset(ImageFormat::Format format, int width, int height, uint32 alignment_boundary); + void Reset(cv::UMat& inputData, ImageFormat::Format format, int width, int height, + uint32_t alignment_boundary); + // Relinquishes ownership of the pixel data. Notice that the unique_ptr // uses a non-standard deleter. std::unique_ptr Release(); @@ -250,6 +258,8 @@ class ImageFrame { int width_step_; std::unique_ptr pixel_data_; + + //OpenClWrapper ocl; }; } // namespace mediapipe diff --git a/mediapipe/framework/formats/image_frame_opencv.cc b/mediapipe/framework/formats/image_frame_opencv.cc index 1ba8c719fa..d79c8372ba 100644 --- a/mediapipe/framework/formats/image_frame_opencv.cc +++ b/mediapipe/framework/formats/image_frame_opencv.cc @@ -11,6 +11,7 @@ // WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. // See the License for the specific language governing permissions and // limitations under the License. +#include #include "mediapipe/framework/formats/image_frame_opencv.h" @@ -87,10 +88,28 @@ cv::Mat MatView(const ImageFrame* image) { CV_MAKETYPE(GetMatType(image->Format()), image->NumberOfChannels()); const size_t steps[] = {static_cast(image->WidthStep()), static_cast(image->ByteDepth())}; + // Use ImageFrame to initialize in-place. ImageFrame still owns memory. return cv::Mat(dims, sizes, type, const_cast(image->PixelData()), steps); } +// UMatUsageFlags +// USAGE_DEFAULT, USAGE_ALLOCATE_HOST_MEMORY, USAGE_ALLOCATE_DEVICE_MEMORY, USAGE_ALLOCATE_SHARED_MEMORY , __UMAT_USAGE_FLAGS_32BIT +cv::UMat MatView(ImageFrame* image, cv::UMatUsageFlags usageFlags) { + //cv::Mat mat = MatView(image); + cv::UMat umat = cv::UMat(usageFlags); + // mat.copyTo(umat); + //*image->MutablePixelData() = static_cast(umat.u->data); + const int type = CV_MAKETYPE(GetMatType(image->Format()), image->NumberOfChannels()); + + //TODO - works only for buffers:CL_MEM_OBJECT_BUFFER == mem_type + cv::ocl::convertFromBuffer(reinterpret_cast(image->MutablePixelData()), static_cast(image->WidthStep()), image->Height(), image->Width(), type, umat); + + // CL_MEM_OBJECT_IMAGE2D == mem_type + //cv::ocl::convertFromImage(reinterpret_cast(image->MutablePixelData()), umat); + return umat; +} + } // namespace formats } // namespace mediapipe diff --git a/mediapipe/framework/formats/image_frame_opencv.h b/mediapipe/framework/formats/image_frame_opencv.h index d197d9a554..c472108242 100644 --- a/mediapipe/framework/formats/image_frame_opencv.h +++ b/mediapipe/framework/formats/image_frame_opencv.h @@ -18,6 +18,7 @@ #include "mediapipe/framework/formats/image_frame.h" #include "mediapipe/framework/port/opencv_core_inc.h" +#include namespace mediapipe { namespace formats { @@ -30,6 +31,10 @@ namespace formats { // even though the returned data is mutable. cv::Mat MatView(const ImageFrame* image); +// UMatUsageFlags +// USAGE_DEFAULT, USAGE_ALLOCATE_HOST_MEMORY, USAGE_ALLOCATE_DEVICE_MEMORY, USAGE_ALLOCATE_SHARED_MEMORY , __UMAT_USAGE_FLAGS_32BIT +cv::UMat MatView(ImageFrame* image, cv::UMatUsageFlags usageFlags); + } // namespace formats } // namespace mediapipe diff --git a/mediapipe/framework/port/BUILD b/mediapipe/framework/port/BUILD index 5894e4715c..ad6f78871c 100644 --- a/mediapipe/framework/port/BUILD +++ b/mediapipe/framework/port/BUILD @@ -242,6 +242,14 @@ cc_library( ], ) +cc_library( + name = "opencv_opencl", + hdrs = ["opencv_opencl.h"], + deps = [ + "//third_party:opencv", + ], +) + cc_library( name = "opencv_imgproc", hdrs = ["opencv_imgproc_inc.h"], diff --git a/mediapipe/framework/port/opencv_opencl.h b/mediapipe/framework/port/opencv_opencl.h new file mode 100644 index 0000000000..c5455b20e7 --- /dev/null +++ b/mediapipe/framework/port/opencv_opencl.h @@ -0,0 +1,20 @@ +// Copyright 2019 The MediaPipe Authors. +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#ifndef MEDIAPIPE_PORT_OPENCV_CORE_INC_H_ +#define MEDIAPIPE_PORT_OPENCV_CORE_INC_H_ + +#include + +#endif // MEDIAPIPE_PORT_OPENCV_CORE_INC_H_ diff --git a/mediapipe/graphs/face_detection/transformation_cpu.pbtxt b/mediapipe/graphs/face_detection/transformation_cpu.pbtxt new file mode 100644 index 0000000000..66d50c7ddf --- /dev/null +++ b/mediapipe/graphs/face_detection/transformation_cpu.pbtxt @@ -0,0 +1,31 @@ +profiler_config { + trace_enabled: true + enable_profiler: true + trace_log_interval_count: 200 + trace_log_path: "/mediapipe/tracing/cpu/" +} +input_stream: "input_video" +output_stream: "output_video" +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE:input_video" + output_stream: "IMAGE:tmp_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 320 + output_height: 320 + } + } +} +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE:tmp_video" + output_stream: "IMAGE:output_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 640 + output_height: 640 + } + } +} + diff --git a/mediapipe/graphs/face_detection/transformation_gpu.pbtxt b/mediapipe/graphs/face_detection/transformation_gpu.pbtxt new file mode 100644 index 0000000000..c85487480c --- /dev/null +++ b/mediapipe/graphs/face_detection/transformation_gpu.pbtxt @@ -0,0 +1,31 @@ +profiler_config { + trace_enabled: true + enable_profiler: true + trace_log_interval_count: 200 + trace_log_path: "/mediapipe/tracing/gpu/" +} +input_stream: "input_video" +output_stream: "output_video" +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE_GPU:input_video" + output_stream: "IMAGE_GPU:tmp_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 320 + output_height: 320 + } + } +} + +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE_GPU:tmp_video" + output_stream: "IMAGE_GPU:output_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 640 + output_height: 640 + } + } +} diff --git a/mediapipe/graphs/object_detection/transformation.pbtxt b/mediapipe/graphs/object_detection/transformation.pbtxt new file mode 100644 index 0000000000..60dd9ae5c4 --- /dev/null +++ b/mediapipe/graphs/object_detection/transformation.pbtxt @@ -0,0 +1,29 @@ +# MediaPipe graph that performs object detection on desktop with TensorFlow Lite +# on CPU. +# Used in the example in +# mediapipe/examples/desktop/object_detection:object_detection_openvino. + +# max_queue_size limits the number of packets enqueued on any input stream +# by throttling inputs to the graph. This makes the graph only process one +# frame per time. +max_queue_size: 1 + +input_stream: "input_video" +output_stream: "output_video" + +# Transforms the input image on CPU to a 320x320 image. To scale the image, by +# default it uses the STRETCH scale mode that maps the entire input image to the +# entire transformed image. As a result, image aspect ratio may be changed and +# objects in the image may be deformed (stretched or squeezed), but the object +# detection model used in this graph is agnostic to that deformation. +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE:input_video" + output_stream: "IMAGE:output_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 320 + output_height: 320 + } + } +} diff --git a/mediapipe/graphs/object_detection/transformation_umat.pbtxt b/mediapipe/graphs/object_detection/transformation_umat.pbtxt new file mode 100644 index 0000000000..17394754c8 --- /dev/null +++ b/mediapipe/graphs/object_detection/transformation_umat.pbtxt @@ -0,0 +1,30 @@ +profiler_config { + trace_enabled: true + enable_profiler: true + trace_log_interval_count: 200 + trace_log_path: "/mediapipe/tracing/" +} +input_stream: "input_video" +output_stream: "output_video" +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE:input_video" + output_stream: "IMAGE:tmp_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 320 + output_height: 320 + } + } +} +node: { + calculator: "ImageTransformationCalculator" + input_stream: "IMAGE:tmp_video" + output_stream: "IMAGE:output_video" + node_options: { + [type.googleapis.com/mediapipe.ImageTransformationCalculatorOptions] { + output_width: 640 + output_height: 640 + } + } +} \ No newline at end of file