diff --git a/commitFile.txt b/commitFile.txt new file mode 100644 index 0000000..95d993f --- /dev/null +++ b/commitFile.txt @@ -0,0 +1,7 @@ +Updating prebuilts and/or headers + +10799426aa623fe55729425a9535e509f56a7800 - nvsample_cudaprocess/customer_functions.h +cf196992f4c45594edc7e57213be89b8db93329a - nvsample_cudaprocess/Makefile +e104b03c418f57b7a5cf24888580dd4867c77525 - nvsample_cudaprocess/nvsample_cudaprocess.cu +651b53fc8c36a0621628150e7fa1a58db8b311c4 - nvsample_cudaprocess/iva_metadata.h +748dc6b13a9cf67aad3caf0fdc97e4e24b34c0bf - nvsample_cudaprocess/nvsample_cudaprocess.h diff --git a/nvsample_cudaprocess/Makefile b/nvsample_cudaprocess/Makefile new file mode 100644 index 0000000..50a83e6 --- /dev/null +++ b/nvsample_cudaprocess/Makefile @@ -0,0 +1,132 @@ +############################################################################### +# +# Copyright (c) 2016-2018, NVIDIA CORPORATION. All rights reserved. +# +# Redistribution and use in source and binary forms, with or without +# modification, are permitted provided that the following conditions +# are met: +# * Redistributions of source code must retain the above copyright +# notice, this list of conditions and the following disclaimer. +# * Redistributions 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. +# * Neither the name of NVIDIA CORPORATION nor the names of its +# contributors may be used to endorse or promote products derived +# from this software without specific prior written permission. +# +# THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. +# +############################################################################### + +# Location of the CUDA Toolkit +CUDA_PATH ?= /usr/local/cuda +INCLUDE_DIR = /usr/include +LIB_DIR = /usr/lib/aarch64-linux-gnu +TEGRA_LIB_DIR = /usr/lib/aarch64-linux-gnu/tegra + +# For hardfp +#LIB_DIR = /usr/lib/arm-linux-gnueabihf +#TEGRA_LIB_DIR = /usr/lib/arm-linux-gnueabihf/tegra + +OSUPPER = $(shell uname -s 2>/dev/null | tr "[:lower:]" "[:upper:]") +OSLOWER = $(shell uname -s 2>/dev/null | tr "[:upper:]" "[:lower:]") + +OS_SIZE = $(shell uname -m | sed -e "s/i.86/32/" -e "s/x86_64/64/" -e "s/armv7l/32/") +OS_ARCH = $(shell uname -m | sed -e "s/i386/i686/") + +GCC ?= g++ +NVCC := $(CUDA_PATH)/bin/nvcc -ccbin $(GCC) + +# internal flags +NVCCFLAGS := --shared +CCFLAGS := -fPIC +LDFLAGS := + +# Extra user flags +EXTRA_NVCCFLAGS ?= +EXTRA_LDFLAGS ?= +EXTRA_CCFLAGS ?= + +override abi := aarch64 +LDFLAGS += --dynamic-linker=/lib/ld-linux-aarch64.so.1 + +# For hardfp +#override abi := gnueabihf +#LDFLAGS += --dynamic-linker=/lib/ld-linux-armhf.so.3 +#CCFLAGS += -mfloat-abi=hard + +ifeq ($(ARMv7),1) +NVCCFLAGS += -target-cpu-arch ARM +ifneq ($(TARGET_FS),) +CCFLAGS += --sysroot=$(TARGET_FS) +LDFLAGS += --sysroot=$(TARGET_FS) +LDFLAGS += -rpath-link=$(TARGET_FS)/lib +LDFLAGS += -rpath-link=$(TARGET_FS)/usr/lib +LDFLAGS += -rpath-link=$(TARGET_FS)/usr/lib/$(abi)-linux-gnu + +# For hardfp +#LDFLAGS += -rpath-link=$(TARGET_FS)/usr/lib/arm-linux-$(abi) + +endif +endif + +# Debug build flags +dbg = 0 +ifeq ($(dbg),1) + NVCCFLAGS += -g -G + TARGET := debug +else + TARGET := release +endif + +ALL_CCFLAGS := +ALL_CCFLAGS += $(NVCCFLAGS) +ALL_CCFLAGS += $(EXTRA_NVCCFLAGS) +ALL_CCFLAGS += $(addprefix -Xcompiler ,$(CCFLAGS)) +ALL_CCFLAGS += $(addprefix -Xcompiler ,$(EXTRA_CCFLAGS)) + +ALL_LDFLAGS := +ALL_LDFLAGS += $(ALL_CCFLAGS) +ALL_LDFLAGS += $(addprefix -Xlinker ,$(LDFLAGS)) +ALL_LDFLAGS += $(addprefix -Xlinker ,$(EXTRA_LDFLAGS)) + +# Common includes and paths for CUDA +INCLUDES := -I./ +LIBRARIES := -L$(LIB_DIR) -lEGL -lGLESv2 +LIBRARIES += -L$(TEGRA_LIB_DIR) -lcuda -lrt + +################################################################################ + +# CUDA code generation flags +GENCODE_SM53 := -gencode arch=compute_53,code=sm_53 +GENCODE_SM62 := -gencode arch=compute_62,code=sm_62 +GENCODE_SM72 := -gencode arch=compute_72,code=sm_72 +GENCODE_SM87 := -gencode arch=compute_87,code=sm_87 +GENCODE_SM_PTX := -gencode arch=compute_87,code=compute_87 +GENCODE_FLAGS ?= $(GENCODE_SM53) $(GENCODE_SM62) $(GENCODE_SM72) $(GENCODE_SM87) $(GENCODE_SM_PTX) + +# Target rules +all: build + +build: libnvsample_cudaprocess.so + +nvsample_cudaprocess.o : nvsample_cudaprocess.cu + $(NVCC) $(INCLUDES) $(ALL_CCFLAGS) $(GENCODE_FLAGS) -o $@ -c $< + +libnvsample_cudaprocess.so : nvsample_cudaprocess.o + $(NVCC) $(ALL_LDFLAGS) $(GENCODE_FLAGS) -o $@ $^ $(LIBRARIES) + +clean: + rm libnvsample_cudaprocess.so nvsample_cudaprocess.o + +clobber: clean diff --git a/nvsample_cudaprocess/customer_functions.h b/nvsample_cudaprocess/customer_functions.h new file mode 100644 index 0000000..f82fc5f --- /dev/null +++ b/nvsample_cudaprocess/customer_functions.h @@ -0,0 +1,104 @@ +/* + * Copyright (c) 2016, NVIDIA CORPORATION. All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions + * are met: + * * Redistributions of source code must retain the above copyright + * notice, this list of conditions and the following disclaimer. + * * Redistributions 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. + * * Neither the name of NVIDIA CORPORATION nor the names of its + * contributors may be used to endorse or promote products derived + * from this software without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. + */ + +#ifndef _CUSTOMER_FUNCTIONS_H_ +#define _CUSTOMER_FUNCTIONS_H_ + +#include + +#if defined(__cplusplus) +extern "C" { +#endif + +typedef enum { + COLOR_FORMAT_Y8 = 0, + COLOR_FORMAT_U8_V8, + COLOR_FORMAT_RGBA, + COLOR_FORMAT_NONE +} ColorFormat; + +typedef struct { + /** + * cuda-process API + * + * @param image : EGL Image to process + * @param userPtr : point to user alloc data, should be free by user + */ + void (*fGPUProcess) (EGLImageKHR image, void ** userPtr); + + /** + * pre-process API + * + * @param sBaseAddr : Mapped Surfaces(YUV) pointers + * @param smemsize : surfaces size array + * @param swidth : surfaces width array + * @param sheight : surfaces height array + * @param spitch : surfaces pitch array + * @param sformat : surfaces format array + * @param nsurfcount : surfaces count + * @param userPtr : point to user alloc data, should be free by user + */ + void (*fPreProcess)(void **sBaseAddr, + unsigned int *smemsize, + unsigned int *swidth, + unsigned int *sheight, + unsigned int *spitch, + ColorFormat *sformat, + unsigned int nsurfcount, + void ** userPtr); + + /** + * post-process API + * + * @param sBaseAddr : Mapped Surfaces(YUV) pointers + * @param smemsize : surfaces size array + * @param swidth : surfaces width array + * @param sheight : surfaces height array + * @param spitch : surfaces pitch array + * @param sformat : surfaces format array + * @param nsurfcount : surfaces count + * @param userPtr : point to user alloc data, should be free by user + */ + void (*fPostProcess)(void **sBaseAddr, + unsigned int *smemsize, + unsigned int *swidth, + unsigned int *sheight, + unsigned int *spitch, + ColorFormat *sformat, + unsigned int nsurfcount, + void ** userPtr); +} CustomerFunction; + +void init (CustomerFunction * pFuncs); +void deinit (void); + +#if defined(__cplusplus) +} +#endif + +#endif//_CUSTOMER_FUNCTIONS_H_ diff --git a/nvsample_cudaprocess/iva_metadata.h b/nvsample_cudaprocess/iva_metadata.h new file mode 100644 index 0000000..d1bcec3 --- /dev/null +++ b/nvsample_cudaprocess/iva_metadata.h @@ -0,0 +1,49 @@ +/* + * Copyright (c) 2016, NVIDIA CORPORATION. All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions + * are met: + * * Redistributions of source code must retain the above copyright + * notice, this list of conditions and the following disclaimer. + * * Redistributions 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. + * * Neither the name of NVIDIA CORPORATION nor the names of its + * contributors may be used to endorse or promote products derived + * from this software without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. + */ + +#ifndef IVA_META_DATA_H +#define IVA_META_DATA_H + +#define NUM_LOCATIONS 8 + +typedef struct _location_t Location; + +struct _location_t +{ + int x1, y1; + int x2, y2; +}; + +typedef struct _bbox_t BBOX; + +struct _bbox_t { + unsigned int framecnt; + unsigned int objectcnt; + Location location_list[NUM_LOCATIONS]; +}; +#endif diff --git a/nvsample_cudaprocess/nvsample_cudaprocess.cu b/nvsample_cudaprocess/nvsample_cudaprocess.cu new file mode 100644 index 0000000..733137f --- /dev/null +++ b/nvsample_cudaprocess/nvsample_cudaprocess.cu @@ -0,0 +1,254 @@ +/* + * Copyright (c) 2016-2018, NVIDIA CORPORATION. All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions + * are met: + * * Redistributions of source code must retain the above copyright + * notice, this list of conditions and the following disclaimer. + * * Redistributions 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. + * * Neither the name of NVIDIA CORPORATION nor the names of its + * contributors may be used to endorse or promote products derived + * from this software without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. + */ + +#include +#include + +#include + +#include "customer_functions.h" +#include "cudaEGL.h" +#include "iva_metadata.h" + +#define BOX_W 32 +#define BOX_H 32 + +#define CORD_X 64 +#define CORD_Y 64 +#define MAX_BUFFERS 30 +static BBOX rect_data[MAX_BUFFERS]; + +/** + * Dummy custom pre-process API implematation. + * It just access mapped surface userspace pointer & + * memset with specific pattern modifying pixel-data in-place. + * + * @param sBaseAddr : Mapped Surfaces pointers + * @param smemsize : surfaces size array + * @param swidth : surfaces width array + * @param sheight : surfaces height array + * @param spitch : surfaces pitch array + * @param nsurfcount : surfaces count + */ +static void +pre_process (void **sBaseAddr, + unsigned int *smemsize, + unsigned int *swidth, + unsigned int *sheight, + unsigned int *spitch, + ColorFormat *sformat, + unsigned int nsurfcount, + void ** usrptr) +{ + /* add your custom pre-process here + we draw a green block for demo */ + int x, y; + char * uv = NULL; + unsigned char * rgba = NULL; + if (sformat[1] == COLOR_FORMAT_U8_V8) { + uv = (char *)sBaseAddr[1]; + for (y = 0; y < BOX_H; ++y) { + for (x = 0; x < BOX_W; ++x) { + uv[y * spitch[1] + 2 * x] = 0; + uv[y * spitch[1] + 2 * x + 1] = 0; + } + } + } else if (sformat[0] == COLOR_FORMAT_RGBA) { + rgba = (unsigned char *)sBaseAddr[0]; + for (y = 0; y < BOX_H*2; y++) { + for (x = 0; x < BOX_W*8; x+=4) { + rgba[x + 0] = 0; + rgba[x + 1] = 0; + rgba[x + 2] = 0; + rgba[x + 3] = 0; + } + rgba+=spitch[0]; + } + } +} + +/** + * Dummy custom post-process API implematation. + * It just access mapped surface userspace pointer & + * memset with specific pattern modifying pixel-data in-place. + * + * @param sBaseAddr : Mapped Surfaces pointers + * @param smemsize : surfaces size array + * @param swidth : surfaces width array + * @param sheight : surfaces height array + * @param spitch : surfaces pitch array + * @param nsurfcount : surfaces count + */ +static void +post_process (void **sBaseAddr, + unsigned int *smemsize, + unsigned int *swidth, + unsigned int *sheight, + unsigned int *spitch, + ColorFormat *sformat, + unsigned int nsurfcount, + void ** usrptr) +{ + /* add your custom post-process here + we draw a green block for demo */ + int x, y; + char * uv = NULL; + int xoffset = (CORD_X * 4); + int yoffset = (CORD_Y * 2); + unsigned char * rgba = NULL; + if (sformat[1] == COLOR_FORMAT_U8_V8) { + uv = (char *)sBaseAddr[1]; + for (y = 0; y < BOX_H; ++y) { + for (x = 0; x < BOX_W; ++x) { + uv[(y + BOX_H * 2) * spitch[1] + 2 * (x + BOX_W * 2)] = 0; + uv[(y + BOX_H * 2) * spitch[1] + 2 * (x + BOX_W * 2) + 1] = 0; + } + } + } else if (sformat[0] == COLOR_FORMAT_RGBA) { + rgba = (unsigned char *)sBaseAddr[0]; + rgba += ((spitch[0] * yoffset) + xoffset); + for (y = 0; y < BOX_H*2; y++) { + for (x = 0; x < BOX_W*8; x+=4) { + rgba[(x + xoffset) + 0] = 0; + rgba[(x + xoffset) + 1] = 0; + rgba[(x + xoffset) + 2] = 0; + rgba[(x + xoffset) + 3] = 0; + } + rgba+=spitch[0]; + } + } +} + +__global__ void addLabelsKernel(int* pDevPtr, int pitch){ + int row = blockIdx.y*blockDim.y + threadIdx.y + BOX_H; + int col = blockIdx.x*blockDim.x + threadIdx.x + BOX_W; + char * pElement = (char*)pDevPtr + row * pitch + col * 2; + pElement[0] = 0; + pElement[1] = 0; + return; +} + +static int addLabels(CUdeviceptr pDevPtr, int pitch){ + dim3 threadsPerBlock(BOX_W, BOX_H); + dim3 blocks(1,1); + addLabelsKernel<<>>((int*)pDevPtr, pitch); + return 0; +} + +static void add_metadata(void ** usrptr) +{ + /* User need to fill rectangle data based on their requirement. + * Here rectangle data is filled for demonstration purpose only */ + + int i; + static int index = 0; + + rect_data[index].framecnt = index; + rect_data[index].objectcnt = index; + + for(i=0; i < NUM_LOCATIONS; i++) + { + rect_data[index].location_list[i].x1 = index; + rect_data[index].location_list[i].x2 = index; + rect_data[index].location_list[i].y1 = index; + rect_data[index].location_list[i].y2 = index; + } + *usrptr = &rect_data[index]; + index++; + if(!(index % MAX_BUFFERS)) + { + index = 0; + } +} + +/** + * Performs CUDA Operations on egl image. + * + * @param image : EGL image + */ +static void +gpu_process (EGLImageKHR image, void ** usrptr) +{ + CUresult status; + CUeglFrame eglFrame; + CUgraphicsResource pResource = NULL; + + cudaFree(0); + status = cuGraphicsEGLRegisterImage(&pResource, image, CU_GRAPHICS_MAP_RESOURCE_FLAGS_NONE); + if (status != CUDA_SUCCESS) { + printf("cuGraphicsEGLRegisterImage failed : %d \n", status); + return; + } + + status = cuGraphicsResourceGetMappedEglFrame( &eglFrame, pResource, 0, 0); + if (status != CUDA_SUCCESS) { + printf ("cuGraphicsSubResourceGetMappedArray failed\n"); + } + + status = cuCtxSynchronize(); + if (status != CUDA_SUCCESS) { + printf ("cuCtxSynchronize failed \n"); + } + + if (eglFrame.frameType == CU_EGL_FRAME_TYPE_PITCH) { + if (eglFrame.eglColorFormat == CU_EGL_COLOR_FORMAT_ABGR) { + /* Rectangle label in plane RGBA, you can replace this with any cuda algorithms */ + addLabels((CUdeviceptr) eglFrame.frame.pPitch[0], eglFrame.pitch); + } else if (eglFrame.eglColorFormat == CU_EGL_COLOR_FORMAT_YUV420_SEMIPLANAR) { + /* Rectangle label in plan UV , you can replace this with any cuda algorithms */ + addLabels((CUdeviceptr) eglFrame.frame.pPitch[1], eglFrame.pitch); + } else + printf ("Invalid eglcolorformat\n"); + } + + add_metadata(usrptr); + + status = cuCtxSynchronize(); + if (status != CUDA_SUCCESS) { + printf ("cuCtxSynchronize failed after memcpy \n"); + } + + status = cuGraphicsUnregisterResource(pResource); + if (status != CUDA_SUCCESS) { + printf("cuGraphicsEGLUnRegisterResource failed: %d \n", status); + } +} + +extern "C" void +init (CustomerFunction * pFuncs) +{ + pFuncs->fPreProcess = pre_process; + pFuncs->fGPUProcess = gpu_process; + pFuncs->fPostProcess = post_process; +} + +extern "C" void +deinit (void) +{ + /* deinitialization */ +} diff --git a/nvsample_cudaprocess/nvsample_cudaprocess.h b/nvsample_cudaprocess/nvsample_cudaprocess.h new file mode 100644 index 0000000..f691c67 --- /dev/null +++ b/nvsample_cudaprocess/nvsample_cudaprocess.h @@ -0,0 +1,36 @@ +/* + * Copyright (c) 2016, NVIDIA CORPORATION. All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions + * are met: + * * Redistributions of source code must retain the above copyright + * notice, this list of conditions and the following disclaimer. + * * Redistributions 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. + * * Neither the name of NVIDIA CORPORATION nor the names of its + * contributors may be used to endorse or promote products derived + * from this software without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. + */ + +#ifndef _CUDA_PROCESS_H_ +#define _CUDA_PROCESS_H_ + +#include "cudaEGL.h" + +extern "C" void Handle_EGLImage (EGLImageKHR image); + +#endif//_CUDA_PROCESS_H_ diff --git a/nvsample_cudaprocess/nvsample_cudaprocess_README.txt b/nvsample_cudaprocess/nvsample_cudaprocess_README.txt new file mode 100644 index 0000000..90894c8 --- /dev/null +++ b/nvsample_cudaprocess/nvsample_cudaprocess_README.txt @@ -0,0 +1,182 @@ +/* + * Copyright (c) 2016-2018, NVIDIA CORPORATION. All rights reserved. + * + * Redistribution and use in source and binary forms, with or without + * modification, are permitted provided that the following conditions + * are met: + * * Redistributions of source code must retain the above copyright + * notice, this list of conditions and the following disclaimer. + * * Redistributions 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. + * * Neither the name of NVIDIA CORPORATION nor the names of its + * contributors may be used to endorse or promote products derived + * from this software without specific prior written permission. + * + * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS ``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 COPYRIGHT OWNER 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. + */ + +a) Install pre-requisites +======================== + +1. Install following packages on Jetson. + sudo apt-get install libegl1-mesa-dev libgles2-mesa-dev libglvnd-dev + +2. Install the NVIDIA(r) CUDA(r) toolkit. (e.g version 8.0) + + Download package from the following website: + https://developer.nvidia.com/embedded/downloads + Ensure that the package name is consistent with the Linux userspace. + + Extract the package with the following command: + $ sudo dpkg -i + + Install the package with the following commands: + $ sudo apt-get update + $ sudo apt-get install cuda-toolkit- + + NOTE: Use proper cuda toolkit version with above installation command. (for e.g. cuda-toolkit-8.0) + +b) Build sample cuda sources +======================== + + $ tar -xpf nvsample_cudaprocess_src.tbz2 + $ cd nvsample_cudaprocess + $ make + $ sudo mv libnvsample_cudaprocess.so /usr/lib/aarch64-linux-gnu/ + +Alternatively, set LD_LIBRARY_PATH as mentioned below instead of moving the library. + $ export LD_LIBRARY_PATH=./ + +c) Run gst-launch-1.0 pipeline +======================== + +Pre-requisite for gstreamer-1.0: Install gstreamer-1.0 plugin using following command on Jetson + + sudo apt-get install gstreamer1.0-tools gstreamer1.0-alsa gstreamer1.0-plugins-base gstreamer1.0-plugins-good gstreamer1.0-plugins-bad gstreamer1.0-plugins-ugly gstreamer1.0-libav libgstreamer1.0-dev libgstreamer-plugins-base1.0-dev libgstreamer-plugins-good1.0-dev + +* Video decode pipeline: + + gst-launch-1.0 filesrc location= ! qtdemux ! h264parse ! omxh264dec ! nvivafilter cuda-process=true customer-lib-name="libnvsample_cudaprocess.so" ! 'video/x-raw(memory:NVMM), format=(string)NV12' ! nvoverlaysink display-id=0 -e + +* Camera capture pipeline: + + gst-launch-1.0 nvcamerasrc fpsRange="30.0 30.0" ! 'video/x-raw(memory:NVMM), width=(int)3840, height=(int)2160, format=(string)I420, framerate=(fraction)30/1' ! nvtee ! nvivafilter cuda-process=true customer-lib-name="libnvsample_cudaprocess.so" ! 'video/x-raw(memory:NVMM), format=(string)NV12' ! nvoverlaysink display-id=0 -e + +NOTE: Make sure the video is larger than 96x96 + +d) Programming Guide +======================== + +1. Sample code in nvsample_cudaprocess_src package + + nvsample_cudaprocess.cu -> sample image pre-/post- processing, CUDA processing functions + pre-process: draw a 32x32 green block start at (0,0) + cuda-process: draw 32x32 green block start at (32,32) + post-process: draw 32x32 green block start at (64,64) + + customer_functions.h -> API definition + +2. Image processing APIs + a. Pre-/Post- processing + i. input parameters: + void ** sBaseAddr : mapped pointers array point to different + plane of image. + + unsigned int * smemsize : actually allocated memory size array for + each image plane, no less than plane + width * height. + + unsigned int * swidth : width array for each image plane + + unsigned int * sheight : height array for each image plane + + unsigned int * spitch : actual line width array in memory for + each image plane, no less than plane + width + + ColorFormat * sformat : color format array, i.e., + * NV12 image will have: + sformat[0] = COLOR_FORMAT_Y8 + sformat[1] = COLOR_FORMAT_U8_V8 + * RGBA image will have: + sformat[0] = COLOR_FORMAT_RGBA + + unsigned int nsurfcount : number of planes of current image type + + void ** userPtr : point to customer allocated buffer in + processing function + + ii. output parameters: + none + + b. CUDA processing + i. input parameters + EGLImageKHR image : Input image data in EGLImage type + void ** userPtr : point to customer allocated buffer in + processing functions + + c. "init" function + This function must be named "init", and accept a pointer to + CustomerFunction structure, which contains 3 function pointers point to + pre-processing, cuda-processing, and post-processing respectively, for + details, please refer to customer_functions.h and nvsample_cudaprocess.cu + + d. "deinit" function + This function must be named "deinit", and is called when the pipeline is + stopping + + e. notes + a customer processing lib: + MUST have an "init" function, which set correspond functions to + nvivafilter plugin; + MAY have a pre-processing function, if not implemented, set to NULL + in "init" function; + MAY have a cuda-processing function, if not imeplemented, set to + NULL in "init" function; + MAY have a post-processing function, if not implemented, set to NULL + in "init" function. + MAY have an "deinit" function if customer functions need to do + deinitialization in stopping the pipeline + +3. Processing Steps + a. nvivafilter plugin input and output + input : (I420, NV12) NVMM buffer, it's NVIDIA's internal frame format, maybe + pitch linear or block linear layout. + output: (NV12, RGBA) NVMM buffer, layout transformed from block linear to pitch linear, + processed result could inplace stored into this buffer. + + b. nvivafilter plugin properties + i. customer-lib-name + string: absolute path and .so lib name to your lib or just the .so + lib name if it is in dynamic lib search path. + + ii. pre-process + bool: dynamically control whether do pre-process if pre-process + function is implemented and set to plugin + + iii. cuda-process + bool: dynamically control whether to do cuda-process if + cuda-process function is implemented and set to plugin + + iv. post-process + bool: dynamically control whether to do post-process if + post-process function is implemented and set to plugin + + c. processing order + customer processing functions will be invoked strictly at following + order if they are implemented and set: + pre-processing -> cuda-processing -> post-processing + plugin property pre-process/cuda-process/post-process can be used for + dynamic enable/disable processing respectively. + diff --git a/push_info.txt b/push_info.txt new file mode 100644 index 0000000..9c21818 --- /dev/null +++ b/push_info.txt @@ -0,0 +1 @@ +jetson_36.2