From 19f12d9c1567662661df76e2e8c761cfc2579aa7 Mon Sep 17 00:00:00 2001 From: perseusdg Date: Mon, 15 Nov 2021 00:08:37 +0530 Subject: [PATCH 1/3] rds_slam -> tensorrt8 port for rds --- include/tkDNN/SegmentationNN.h | 241 +++++++++++++++++++++++++++++++++ 1 file changed, 241 insertions(+) diff --git a/include/tkDNN/SegmentationNN.h b/include/tkDNN/SegmentationNN.h index 403bc28..fff27a8 100644 --- a/include/tkDNN/SegmentationNN.h +++ b/include/tkDNN/SegmentationNN.h @@ -18,6 +18,7 @@ #include "tkdnn.h" #include "NetworkViz.h" #include "kernelsThrust.h" +#define SLAM_MODE namespace tk { namespace dnn { @@ -99,6 +100,62 @@ class SegmentationNN { * * @param bi batch index */ + + #ifdef SLAM_MODE + cv::Mat postprocess(const int bi=0,bool apply_colormap=true){ + cv::Mat maskMatrix; + dnnType *rt_out = (dnnType *)netRT->buffersRT[1]+ netRT->buffersDIM[1].tot()*bi; + + dataDim_t odim = netRT->output_dim; + + matrixTranspose(cublasHandle, rt_out, tmpInputData_d, odim.c, odim.w*odim.h); + maxElem(tmpInputData_d, tmpOutData_d, odim.c, odim.h, odim.w); + checkCuda(cudaMemcpy(tmpOutData_h, tmpOutData_d, odim.w*odim.h * sizeof(float), cudaMemcpyDeviceToHost)); + + + + dataDim_t vdim = odim; + vdim.c = 1; + dnnType *dataTemp = nullptr; + if(isCudaPointer(tmpOutData_h)) + { + dataTemp = new dnnType[vdim.tot()]; + checkCuda(cudaMemcpy(dataTemp,tmpOutData_h,vdim.tot()*sizeof(dnnType),cudaMemcpyDeviceToHost)); + } + else + { + dataTemp = tmpOutData_h; + } + for(int i =0;iinput_dim.h, netRT->input_dim.w, 0, classes, classes); + else{ + cv::Mat colored_fp32 (cv::Size(odim.w, odim.h),CV_32FC1, dataTemp); + colored_fp32.convertTo(colored, CV_8UC1); + } + + int max_dim = (originalSize[bi].width > originalSize[bi].height) ? originalSize[bi].width : originalSize[bi].height; + resize(colored, colored, cv::Size(max_dim, max_dim)); + int top, bottom, left, right; + computeBorders(originalSize[bi].width, originalSize[bi].height, top, bottom, left, right); + cv::Rect roi(left,top,originalSize[bi].width, originalSize[bi].height); + cv::Mat or_size (colored, roi); + segmented[bi] = or_size; + + if(isCudaPointer(tmpOutData_h)) + { + delete [] dataTemp; + } + + return maskMatrix; + + } + #elif + void postprocess(const int bi=0, bool appy_colormap = true) { dnnType *rt_out = (dnnType *)netRT->buffersRT[1]+ netRT->buffersDIM[1].tot()*bi; @@ -128,6 +185,7 @@ class SegmentationNN { cv::Mat or_size (colored, roi); segmented[bi] = or_size; }; + #endif public: int classes = 0; @@ -237,6 +295,184 @@ class SegmentationNN { } } + #ifdef SLAM_MODE + cv::Mat updateOriginal(cv::Mat frame,bool apply_colormap=true){ + std::vector splitted_frames; + cv::Mat maskMatrix; + int H, W, net_H, net_W; + int top = 0, bottom = 0, left = 0, right = 0; + std::vector> pos; + + { + TKDNN_TSTART + cv::Size original_size = frame.size(); + + frame.convertTo(frame, CV_32FC3, 1 / 255.0, 0); + H = frame.rows; + W = frame.cols; + net_H = netRT->input_dim.h; + net_W = netRT->input_dim.w; + + cv::Mat frame_cropped; + + if( H <= net_H && W <= net_W ){ // smaller size wrt network + top = (net_H - H)/2; + bottom = net_H - H - top ; + left = (net_W - W)/2; + right = net_W - W - left ; + cv::copyMakeBorder(frame, frame_cropped, top, bottom, left, right, cv::BORDER_CONSTANT, cv::Scalar(0,0,0) ); + splitted_frames.push_back(frame_cropped); + } + else{ //bigger size wrt network + + + if(H < net_H || W < net_W){ + if(H < net_H){ + top = (net_H - H)/2; + bottom = net_H - H - top ; + } + else{ + left = (net_W - W)/2; + right = net_W - W - left ; + } + cv::copyMakeBorder(frame, frame_cropped, top, bottom, left, right, cv::BORDER_CONSTANT, cv::Scalar(0,0,0)); + } + + for(int x=0; x+net_W<=W ;){ + for(int y=0; y+net_H <=H ; ){ + cv::Rect roi(x, y, net_W, net_H); + cv::Mat image_roi = frame(roi); + splitted_frames.push_back(image_roi); + pos.push_back(std::make_pair(x,y)); + + y += net_H; + if(y == H) + break; + if(y + net_H > H) y = H - net_H; + } + x += net_W; + if(x == W) + break; + if(x + net_W > W) x = W - net_W; + } + } + + tk::dnn::dataDim_t idim = netRT->input_dim; + + if(splitted_frames.size()> nBatches) + FatalError(std::to_string(splitted_frames.size()) + " min batches required"); + + for(int bi=0; bistream)); + normalize(input_d + idim.tot()*bi, idim.c, idim.h, idim.w, mean_d, stddev_d); + } + TKDNN_TSTOP + stats_pre.push_back(t_ns); + } + + tk::dnn::dataDim_t dim = netRT->input_dim; + dim.n = splitted_frames.size(); + { + if(TKDNN_VERBOSE) dim.print(); + TKDNN_TSTART + netRT->infer(dim, input_d); + TKDNN_TSTOP + if(TKDNN_VERBOSE) dim.print(); + stats.push_back(t_ns); + } + + dataDim_t odim = netRT->output_dim; + + std::vector out_img; + std::vector out_mask; + + { + TKDNN_TSTART + + for(int bi=0; bibuffersRT[1]+ netRT->buffersDIM[1].tot()*bi; + + matrixTranspose(cublasHandle, rt_out, tmpInputData_d, odim.c, odim.w*odim.h); + maxElem(tmpInputData_d, tmpOutData_d, odim.c, odim.h, odim.w); + checkCuda(cudaMemcpy(tmpOutData_h, tmpOutData_d, odim.w*odim.h * sizeof(float), cudaMemcpyDeviceToHost)); + + dataDim_t vdim = odim; + vdim.c = 1; + dnnType *dataTemp = nullptr; + if(isCudaPointer(tmpOutData_h)) + { + dataTemp = new dnnType[vdim.tot()]; + checkCuda(cudaMemcpy(dataTemp,tmpOutData_h,vdim.tot()*sizeof(dnnType),cudaMemcpyDeviceToHost)); + } + else + { + dataTemp = tmpOutData_h; + } + + cv::Mat colored; + for(int i=0;iinput_dim.h, netRT->input_dim.w, 0, classes, classes); + else{ + cv::Mat colored_fp32 (cv::Size(odim.w, odim.h),CV_32FC1, tmpOutData_h); + colored_fp32.convertTo(colored, CV_8UC1); + } + out_img.push_back(colored); + if(isCudaPointer(tmpOutData_h)) + { + delete [] dataTemp; + } + } + + cv::Mat tempMask(frame.size(), out_mask[0].type()); + cv::Mat seg(frame.size(), out_img[0].type()); + if(out_img.size() == 1) + { + cv::Rect roi(left, top, W, H); + seg = out_img[0](roi); + tempMask = out_mask[0](roi); + } + else{ + int bi=0; + + if(top == 0 && left == 0){ + + for(int i=0; i splitted_frames; @@ -385,6 +621,11 @@ class SegmentationNN { stats_post.push_back(t_ns); } } + #endif + + + + /** * Method to draw boundixg boxes and labels on a frame. From 4d8f99b44143c8c140e98a2efc4b04e198f49d85 Mon Sep 17 00:00:00 2001 From: perseusdg Date: Sun, 9 Jan 2022 19:24:38 +0530 Subject: [PATCH 2/3] depth->tensorrt8 patches --- CMakeLists.txt | 5 +- include/tkDNN/NetworkRT.h | 2 + include/tkDNN/pluginsRT/ReflectionPadding.h | 2 + src/NetworkRT.cpp | 55 +++++++++++++++------ 4 files changed, 46 insertions(+), 18 deletions(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 7cc9e33..d919b46 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -85,12 +85,14 @@ endif() find_package(CUDNN REQUIRED) include_directories(${CUDNN_INCLUDE_DIR}) +find_package(yaml-cpp REQUIRED) + # compile file(GLOB tkdnn_CUSRC "src/kernels/*.cu" "src/sorting.cu" "src/pluginsRT/*.cpp") cuda_include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS} ${CUDNN_INCLUDE_DIRS}) cuda_add_library(kernels SHARED ${tkdnn_CUSRC}) -target_link_libraries(kernels ${CUDA_CUBLAS_LIBRARIES} ${CUDA_LIBRARIES} ${CUDNN_LIBRARIES}) +target_link_libraries(kernels ${CUDA_CUBLAS_LIBRARIES} ${CUDA_LIBRARIES} ${CUDNN_LIBRARIES} yaml-cpp) @@ -120,7 +122,6 @@ endif() # endif() # gives problems in cross-compiling, probably malformed cmake config -find_package(yaml-cpp REQUIRED) #------------------------------------------------------------------------------- # Build Libraries diff --git a/include/tkDNN/NetworkRT.h b/include/tkDNN/NetworkRT.h index 571d127..b859074 100644 --- a/include/tkDNN/NetworkRT.h +++ b/include/tkDNN/NetworkRT.h @@ -23,6 +23,8 @@ #include #include #include +#include +#include diff --git a/include/tkDNN/pluginsRT/ReflectionPadding.h b/include/tkDNN/pluginsRT/ReflectionPadding.h index 894ed98..7b13710 100644 --- a/include/tkDNN/pluginsRT/ReflectionPadding.h +++ b/include/tkDNN/pluginsRT/ReflectionPadding.h @@ -94,6 +94,8 @@ namespace nvinfer1{ static std::vector mPluginAttributes; std::string mPluginNamespace; }; + + REGISTER_TENSORRT_PLUGIN(ReflectionPaddingRTPluginCreator); }; #endif diff --git a/src/NetworkRT.cpp b/src/NetworkRT.cpp index 6192006..3f086bd 100644 --- a/src/NetworkRT.cpp +++ b/src/NetworkRT.cpp @@ -474,23 +474,46 @@ ILayer* NetworkRT::convert_layer(ITensor *input,Padding *l){ #else //todo add PADDING_MODE_CONSTANT AND PADDING_MODE_ZERO for tensorrt versions < 8.2 if(l->padding_mode == PADDING_MODE_REFLECTION){ - auto creator = getPluginRegistry()->getPluginCreator("ReflectionPaddingRT_tkDNN","1"); - std::vector mPluginAttributes; - PluginFieldCollection mFC{}; - mPluginAttributes.emplace_back(PluginField("padH",&l->paddingH,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("padW",&l->paddingW,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("inputH",&l->input_dim.h,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("inputW",&l->input_dim.w,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("outputH",&l->output_dim.h,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("outputW",&l->output_dim.w,PluginFieldType::kINT32,1)); - mPluginAttributes.emplace_back(PluginField("n",&l->input_dim.n,PluginFieldType::kINT32,1)); - mFC.nbFields = mPluginAttributes.size(); - mFC.fields = mPluginAttributes.data(); - auto *plugin = creator->createPlugin(l->getLayerName().c_str(),&mFC); + auto creator = getPluginRegistry()->getPluginCreator("ReflectionPaddingRT_tkDNN","1"); + std::vector mPluginAttributes; + PluginFieldCollection mFC{}; + mPluginAttributes.emplace_back(PluginField("padH",&l->paddingH,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("padW",&l->paddingW,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("inputH",&l->input_dim.h,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("inputW",&l->input_dim.w,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("outputH",&l->output_dim.h,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("outputW",&l->output_dim.w,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("n",&l->input_dim.n,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("c",&l->input_dim.c,PluginFieldType::kINT32,1)); + mFC.nbFields = mPluginAttributes.size(); + mFC.fields = mPluginAttributes.data(); + auto *plugin = creator->createPlugin(l->getLayerName().c_str(),&mFC); + auto *lRT = networkRT->addPluginV2(&input, 1, *plugin); + checkNULL(lRT); + return lRT; + }else if(l->padding_mode == PADDING_MODE_CONSTANT || l->padding_mode == PADDING_MODE_ZERO){ + auto creator = getPluginRegistry()->getPluginCreator("ConstantPaddingRT_tkDNN","1"); + std::vector mPluginAttributes; + PluginFieldCollection mFC{}; + mPluginAttributes.emplace_back(PluginField("padH",&l->paddingH,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("padW",&l->paddingW,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("inputH",&l->input_dim.h,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("inputW",&l->input_dim.w,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("outputH",&l->output_dim.h,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("outputW",&l->output_dim.w,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("n",&l->input_dim.n,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("c",&l->input_dim.c,PluginFieldType::kINT32,1)); + mPluginAttributes.emplace_back(PluginField("constant",&l->constant,PluginFieldType::kFLOAT32,1)); + mFC.nbFields = mPluginAttributes.size(); + mFC.fields = mPluginAttributes.data(); + auto *plugin = creator->createPlugin(l->getLayerName().c_str(),&mFC); + auto *lRT = networkRT->addPluginV2(&input,1,*plugin); + checkNULL(lRT); + return lRT; } - auto *lRT = networkRT->addPluginV2(&input, 1, *plugin); - checkNULL(lRT); - return lRT; + + return nullptr; + #endif } From 30239266954c59567c4c72404eabeb0a9b176bb2 Mon Sep 17 00:00:00 2001 From: perseusdg Date: Sun, 9 Jan 2022 22:17:43 +0530 Subject: [PATCH 3/3] fix constantpadding commit --- include/tkDNN/pluginsRT/ConstantPaddingRT.h | 109 +++++++++++ src/pluginsRT/ConstantPaddingRT.cpp | 201 ++++++++++++++++++++ 2 files changed, 310 insertions(+) create mode 100644 include/tkDNN/pluginsRT/ConstantPaddingRT.h create mode 100644 src/pluginsRT/ConstantPaddingRT.cpp diff --git a/include/tkDNN/pluginsRT/ConstantPaddingRT.h b/include/tkDNN/pluginsRT/ConstantPaddingRT.h new file mode 100644 index 0000000..15f4c0d --- /dev/null +++ b/include/tkDNN/pluginsRT/ConstantPaddingRT.h @@ -0,0 +1,109 @@ +// +// Created by perseusdg on 1/7/22. +// + +#ifndef _CONSTANTPADDINGRT_PLUGIN_H +#define _CONSTANTPADDINGRT_PLUGIN_H + +#include +#include +#include +#include +#include + +namespace nvinfer1{ + class ConstantPaddingRT : public IPluginV2Ext { + public: + ConstantPaddingRT(int32_t padH,int32_t padW,int32_t n,int32_t c,int32_t i_h,int32_t i_w,int32_t o_h,int32_t o_w,float constant); + + ConstantPaddingRT(const void *data,size_t length); + + ~ConstantPaddingRT(); + + int getNbOutputs() const NOEXCEPT override; + + Dims getOutputDimensions(int index, const Dims *inputs, int nbInputDims) NOEXCEPT override ; + + int initialize() NOEXCEPT override ; + + void terminate() NOEXCEPT override ; + + size_t getWorkspaceSize(int maxBatchSize) const NOEXCEPT override ; + + +#if NV_TENSORRT_MAJOR > 7 + int enqueue(int batchSize, const void *const *inputs, void *const *outputs, void *workspace, cudaStream_t stream) NOEXCEPT override ; +#elif NV_TENSORRT_MAJOR <= 7 + int32_t enqueue (int32_t batchSize, const void *const *inputs, void **outputs, void *workspace, cudaStream_t stream) override; +#endif + + size_t getSerializationSize() const NOEXCEPT override ; + + void serialize(void *buffer) const NOEXCEPT override ; + + void destroy() NOEXCEPT override ; + + const char *getPluginType() const NOEXCEPT override ; + + const char *getPluginVersion() const NOEXCEPT override; + + const char *getPluginNamespace() const NOEXCEPT override ; + + void setPluginNamespace(const char *pluginNamespace) NOEXCEPT override ; + + IPluginV2Ext *clone() const NOEXCEPT override ; + + DataType getOutputDataType(int index, const nvinfer1::DataType* inputTypes, int nbInputs) const NOEXCEPT override; + + void attachToContext(cudnnContext* cudnnContext, cublasContext* cublasContext, IGpuAllocator* gpuAllocator) NOEXCEPT override; + + bool isOutputBroadcastAcrossBatch(int outputIndex, const bool* inputIsBroadcasted, int nbInputs) const NOEXCEPT override; + + bool canBroadcastInputAcrossBatch(int inputIndex) const NOEXCEPT override; + + void configurePlugin (Dims const *inputDims, int32_t nbInputs, Dims const *outputDims, + int32_t nbOutputs, DataType const *inputTypes, DataType const *outputTypes, + bool const *inputIsBroadcast, bool const *outputIsBroadcast, PluginFormat floatFormat, + int32_t maxBatchSize) NOEXCEPT override; + + void detachFromContext() NOEXCEPT override; + + bool supportsFormat (DataType type, PluginFormat format) const NOEXCEPT override; + + int32_t i_h,i_w,o_h,o_w,n,c,padH,padW; + float constant; + private: + std::string mPluginNamespace; + + }; + + class ConstantPaddingRTPluginCreator : public IPluginCreator { + public: + ConstantPaddingRTPluginCreator(); + + void setPluginNamespace(const char* pluginNamespace) NOEXCEPT override; + + const char *getPluginNamespace() const NOEXCEPT override; + + IPluginV2Ext *deserializePlugin(const char *name, const void *serialData, size_t serialLength) NOEXCEPT override ; + + IPluginV2Ext *createPlugin(const char *name, const PluginFieldCollection *fc) NOEXCEPT override ; + + const char *getPluginName() const NOEXCEPT override ; + + const char *getPluginVersion() const NOEXCEPT override; + + const PluginFieldCollection *getFieldNames() NOEXCEPT override ; + + private: + static PluginFieldCollection mFC; + static std::vector mPluginAttributes; + std::string mPluginNamespace; + + }; + + REGISTER_TENSORRT_PLUGIN(ConstantPaddingRTPluginCreator); +}; + + +#endif //TKDNN_CONSTANTPADDINGRT_H diff --git a/src/pluginsRT/ConstantPaddingRT.cpp b/src/pluginsRT/ConstantPaddingRT.cpp new file mode 100644 index 0000000..ac37e9d --- /dev/null +++ b/src/pluginsRT/ConstantPaddingRT.cpp @@ -0,0 +1,201 @@ +#include + +using namespace nvinfer1; + +std::vector ConstantPaddingRTPluginCreator::mPluginAttributes; +PluginFieldCollection ConstantPaddingRTPluginCreator::mFC{}; + +static const char* CONSTANTPADDINGRT_PLUGIN_VERSION{"1"}; +static const char* CONSTANTPADDINGRT_PLUGIN_NAME{"ConstantPaddingRT_tkDNN"}; + +ConstantPaddingRT::ConstantPaddingRT(int32_t padH, int32_t padW, int32_t n, int32_t c, int32_t i_h, int32_t i_w, + int32_t o_h, int32_t o_w, float constant) { + this->padH = padH; + this->padW = padW; + this->n = n; + this->c = c; + this->i_h = i_h; + this->i_w = i_w; + this->o_h = o_h; + this->o_w = o_w; + this->constant = constant; + +} + +ConstantPaddingRT::ConstantPaddingRT(const void *data, size_t length) { + const char* buf = reinterpret_cast(data),*bufcheck=buf; + padH = readBUF(buf); + padW = readBUF(buf); + i_h = readBUF(buf); + i_w = readBUF(buf); + o_h = readBUF(buf); + o_w = readBUF(buf); + n = readBUF(buf); + c = readBUF(buf); + constant = readBUF(buf); + assert(buf = bufcheck + length); +} + +ConstantPaddingRT::~ConstantPaddingRT() {} + +int ConstantPaddingRT::getNbOutputs() const NOEXCEPT{ + return 1; +} + +Dims ConstantPaddingRT::getOutputDimensions(int index, const Dims *inputs, int nbInputDims) NOEXCEPT { + return Dims3{c,o_h,o_w}; +} + +int ConstantPaddingRT::initialize() NOEXCEPT { + return 0; +} + +void ConstantPaddingRT::terminate() NOEXCEPT { + +} + +size_t ConstantPaddingRT::getWorkspaceSize(int maxBatchSize) const NOEXCEPT { + return 0; +} + +#if NV_TENSORRT_MAJOR > 7 +int ConstantPaddingRT::enqueue(int batchSize, const void *const *inputs, void *const *outputs, void *workspace, cudaStream_t stream) NOEXCEPT { + dnnType* srcData = (dnnType*)reinterpret_cast(inputs[0]); + dnnType* dstData = reinterpret_cast(outputs[0]); + constant_pad2d_forward(srcData,dstData,i_h,i_w,o_h,o_w,c,n,padH,padW,constant,stream); + return 0; +} +#elif NV_TENSORRT_MAJOR <= 7 + int32_t enqueue (int32_t batchSize, const void *const *inputs, void **outputs, void *workspace, cudaStream_t stream) { + dnnType* srcData = (dnnType*)reinterpret_cast(inputs[0]); + dnnType* dstData = reinterpret_cast(outputs[0]); + constant_pad2d_forward(srcData,dstData,i_h,i_w,o_h,o_w,c,n,padH,padW,constant,stream); + return 0; +} +#endif + +size_t ConstantPaddingRT::getSerializationSize() const NOEXCEPT { + return (8*sizeof(int32_t) + 1*sizeof(float)); +} + +void ConstantPaddingRT::serialize(void *buffer) const NOEXCEPT { + char *buf = reinterpret_cast(buffer),*a=buf; + writeBUF(buf,padH); + writeBUF(buf,padW); + writeBUF(buf,i_h); + writeBUF(buf,i_w); + writeBUF(buf,o_h); + writeBUF(buf,o_w); + writeBUF(buf,n); + writeBUF(buf,c); + writeBUF(buf,constant); +} + +void ConstantPaddingRT::destroy() NOEXCEPT { + delete this; +} + +const char* ConstantPaddingRT::getPluginType() const NOEXCEPT { + return CONSTANTPADDINGRT_PLUGIN_NAME; +} + +const char* ConstantPaddingRT::getPluginVersion() const NOEXCEPT { + return CONSTANTPADDINGRT_PLUGIN_VERSION; +} + +const char* ConstantPaddingRT::getPluginNamespace() const NOEXCEPT { + return mPluginNamespace.c_str(); +} + +void ConstantPaddingRT::setPluginNamespace(const char *pluginNamespace) NOEXCEPT { + mPluginNamespace = pluginNamespace; +} + +IPluginV2Ext *ConstantPaddingRT::clone() const NOEXCEPT { + auto *p = new ConstantPaddingRT(padH,padW,n,c,i_h,i_w,o_h,o_w,constant); + p->setPluginNamespace(mPluginNamespace.c_str()); + return p; +} + +DataType ConstantPaddingRT::getOutputDataType(int index, const nvinfer1::DataType *inputTypes, + int nbInputs) const NOEXCEPT { + return DataType::kFLOAT; +} + +void ConstantPaddingRT::attachToContext(cudnnContext *cudnnContext, cublasContext *cublasContext, + IGpuAllocator *gpuAllocator) NOEXCEPT { + +} + +bool ConstantPaddingRT::isOutputBroadcastAcrossBatch(int outputIndex, const bool *inputIsBroadcasted, + int nbInputs) const NOEXCEPT { + return false; +} + +bool ConstantPaddingRT::canBroadcastInputAcrossBatch(int inputIndex) const NOEXCEPT { + return false; +} + +void ConstantPaddingRT::configurePlugin(const Dims *inputDims, int32_t nbInputs, const Dims *outputDims, + int32_t nbOutputs, const DataType *inputTypes, const DataType *outputTypes, + const bool *inputIsBroadcast, const bool *outputIsBroadcast, + PluginFormat floatFormat, int32_t maxBatchSize) NOEXCEPT { + +} + +void ConstantPaddingRT::detachFromContext() NOEXCEPT { + +} + +bool ConstantPaddingRT::supportsFormat(DataType type, PluginFormat format) const NOEXCEPT { + return (type == DataType::kFLOAT && format == PluginFormat::kLINEAR); +} + +ConstantPaddingRTPluginCreator::ConstantPaddingRTPluginCreator() { + mPluginAttributes.clear(); + mFC.nbFields = mPluginAttributes.size(); + mFC.fields = mPluginAttributes.data(); +} + +void ConstantPaddingRTPluginCreator::setPluginNamespace(const char *pluginNamespace) NOEXCEPT { + mPluginNamespace = pluginNamespace; +} + +const char *ConstantPaddingRTPluginCreator::getPluginNamespace() const NOEXCEPT { + return mPluginNamespace.c_str(); +} + +IPluginV2Ext *ConstantPaddingRTPluginCreator::deserializePlugin(const char *name, const void *serialData, + size_t serialLength) NOEXCEPT { + auto *pluginObj = new ConstantPaddingRT(serialData,serialLength); + pluginObj->setPluginNamespace(mPluginNamespace.c_str()); + return pluginObj; +} + +IPluginV2Ext *ConstantPaddingRTPluginCreator::createPlugin(const char *name, const PluginFieldCollection *fc) NOEXCEPT { + const PluginField *fields = fc->fields; + int padH = *(static_cast(fields[0].data)); + int padW = *(static_cast(fields[1].data)); + int inputH = *(static_cast(fields[2].data)); + int inputW = *(static_cast(fields[3].data)); + int outputH = *(static_cast(fields[4].data)); + int outputW = *(static_cast(fields[5].data)); + int n = *(static_cast(fields[6].data)); + int c = *(static_cast(fields[7].data)); + float constant = *(static_cast(fields[8].data)); + auto *pluginObj = new ConstantPaddingRT(padH,padW,n,c,inputH,inputW,outputH,outputW,constant); + pluginObj->setPluginNamespace(mPluginNamespace.c_str()); + return pluginObj; +} + +const char *ConstantPaddingRTPluginCreator::getPluginName() const NOEXCEPT { + return CONSTANTPADDINGRT_PLUGIN_NAME; +} + +const char *ConstantPaddingRTPluginCreator::getPluginVersion() const NOEXCEPT { + return CONSTANTPADDINGRT_PLUGIN_VERSION; +} + +const PluginFieldCollection *ConstantPaddingRTPluginCreator::getFieldNames() NOEXCEPT { + return &mFC; +}