From 2e8d0b1002d552aa0b430dae89408ce6ae40580f Mon Sep 17 00:00:00 2001 From: Francesco Gatti Date: Sun, 23 Dec 2018 16:20:17 +0100 Subject: [PATCH] yolo3 86 route error --- include/NetworkRT.h | 4 +++- src/NetworkRT.cpp | 26 ++++++++++++++++--------- src/Shortcut.cpp | 5 +++++ tests/yolo3_berkeley/yolo3_berkeley.cpp | 18 ++++------------- 4 files changed, 29 insertions(+), 24 deletions(-) diff --git a/include/NetworkRT.h b/include/NetworkRT.h index c1e8d20..5a11a32 100644 --- a/include/NetworkRT.h +++ b/include/NetworkRT.h @@ -18,7 +18,9 @@ public: nvinfer1::ICudaEngine *engineRT; nvinfer1::IExecutionContext *contextRT; - void* buffersRT[2]; + + const static int MAX_BUFFERS_RT = 10; + void* buffersRT[MAX_BUFFERS_RT]; int buf_input_idx, buf_output_idx; dataDim_t input_dim, output_dim; diff --git a/src/NetworkRT.cpp b/src/NetworkRT.cpp index 522f05f..c14fa44 100644 --- a/src/NetworkRT.cpp +++ b/src/NetworkRT.cpp @@ -77,6 +77,9 @@ NetworkRT::NetworkRT(Network *net, const char *name) { Ilay->setName( (l->getLayerName() + std::to_string(i)).c_str() ); input = Ilay->getOutput(0); + if(l->getLayerType() == LAYER_YOLO) + networkRT->markOutput(*input); + tensors[l] = input; } if(input == NULL) @@ -99,9 +102,9 @@ NetworkRT::NetworkRT(Network *net, const char *name) { contextRT = engineRT->createExecutionContext(); // input and output buffer pointers that we pass to the engine - the engine requires exactly IEngine::getNbBindings(), - // of these, but in this case we know that there is exactly one input and one output. - if(engineRT->getNbBindings() != 2) - FatalError("Incorrect buffers number"); + std::cout<<"Input/outputs numbers: "<getNbBindings()<<"\n"; + if(engineRT->getNbBindings() > MAX_BUFFERS_RT) + FatalError("over RT buffer array size"); // In order to bind the buffers, we need to know the names of the input and output tensors. // note that indices are guaranteed to be less than IEngine::getNbBindings() @@ -122,10 +125,13 @@ NetworkRT::NetworkRT(Network *net, const char *name) { output_dim.c = oDim.d[0]; output_dim.h = oDim.d[1]; output_dim.w = oDim.d[2]; + output_dim.print(); // create GPU buffers and a stream - checkCuda(cudaMalloc(&buffersRT[buf_input_idx], input_dim.tot()*sizeof(dnnType))); - checkCuda(cudaMalloc(&buffersRT[buf_output_idx], output_dim.tot()*sizeof(dnnType))); + for(int i=0; igetNbBindings(); i++) { + Dims dim = engineRT->getBindingDimensions(i); + checkCuda(cudaMalloc(&buffersRT[i], dim.d[0]*dim.d[1]*dim.d[2]*sizeof(dnnType))); + } checkCuda(cudaMalloc(&output, output_dim.tot()*sizeof(dnnType))); checkCuda(cudaStreamCreate(&stream)); } @@ -136,9 +142,9 @@ NetworkRT::~NetworkRT() { dnnType* NetworkRT::infer(dataDim_t &dim, dnnType* data) { - checkCuda(cudaMemcpyAsync(buffersRT[buf_input_idx], data, input_dim.tot()*sizeof(float), cudaMemcpyDeviceToDevice, stream)); + checkCuda(cudaMemcpyAsync(buffersRT[buf_input_idx], data, input_dim.tot()*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream)); contextRT->enqueue(1, buffersRT, stream, nullptr); - checkCuda(cudaMemcpyAsync(output, buffersRT[buf_output_idx], output_dim.tot()*sizeof(float), cudaMemcpyDeviceToDevice, stream)); + checkCuda(cudaMemcpyAsync(output, buffersRT[buf_output_idx], output_dim.tot()*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream)); cudaStreamSynchronize(stream); dim = output_dim; @@ -301,7 +307,7 @@ ILayer* NetworkRT::convert_layer(ITensor *input, Softmax *l) { ILayer* NetworkRT::convert_layer(ITensor *input, Route *l) { //std::cout<<"convert route\n"; - ITensor *tens[256]; + ITensor **tens = new ITensor*[l->layers_n]; for(int i=0; ilayers_n; i++) tens[i] = tensors[l->layers[i]]; IConcatenationLayer *lRT = networkRT->addConcatenation(tens, l->layers_n); @@ -337,7 +343,9 @@ ILayer* NetworkRT::convert_layer(ITensor *input, Shortcut *l) { ITensor *back_tens = tensors[l->backLayer]; IPlugin *plugin = new ShortcutRT(); - ITensor *inputs[2] = { input, back_tens }; + ITensor **inputs = new ITensor*[2]; + inputs[0] = input; + inputs[1] = back_tens; IPluginLayer *lRT = networkRT->addPlugin(inputs, 2, *plugin); checkNULL(lRT); return lRT; diff --git a/src/Shortcut.cpp b/src/Shortcut.cpp index 1d817fe..bcd2a00 100644 --- a/src/Shortcut.cpp +++ b/src/Shortcut.cpp @@ -9,6 +9,11 @@ Shortcut::Shortcut(Network *net, Layer *backLayer) : Layer(net) { this->backLayer = backLayer; checkCuda( cudaMalloc(&dstData, output_dim.tot()*sizeof(dnnType)) ); + + if( backLayer->output_dim.c != input_dim.c || + backLayer->output_dim.w != input_dim.w || + backLayer->output_dim.h != input_dim.h ) + FatalError("Shortcut dim missmatch"); } Shortcut::~Shortcut() { diff --git a/tests/yolo3_berkeley/yolo3_berkeley.cpp b/tests/yolo3_berkeley/yolo3_berkeley.cpp index 2ea1d58..b08836d 100644 --- a/tests/yolo3_berkeley/yolo3_berkeley.cpp +++ b/tests/yolo3_berkeley/yolo3_berkeley.cpp @@ -231,7 +231,6 @@ int main() { tk::dnn::Activation a78 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c79 (&net, 512, 1, 1, 1, 1, 0, 0, c79_bin, true); tk::dnn::Activation a79 (&net, tk::dnn::ACTIVATION_LEAKY); -/* tk::dnn::Conv2d c80 (&net,1024, 3, 3, 1, 1, 1, 1, c80_bin, true); tk::dnn::Activation a80 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c81 (&net, 45, 1, 1, 1, 1, 0, 0, c81_bin, false); @@ -239,13 +238,12 @@ int main() { tk::dnn::Layer *m83_layers[1] = { &a79 }; tk::dnn::Route m83 (&net, m83_layers, 1); -*/ tk::dnn::Conv2d c84 (&net, 256, 1, 1, 1, 1, 0, 0, c84_bin, true); tk::dnn::Activation a84 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Upsample u85 (&net, 2); -// tk::dnn::Layer *m86_layers[2] = { &u85, &s61 }; -// tk::dnn::Route m86 (&net, m86_layers, 2); + tk::dnn::Layer *m86_layers[2] = { &u85, &s61 }; + tk::dnn::Route m86 (&net, m86_layers, 1); // ROUTE ERROR IN RT INFERENCE tk::dnn::Conv2d c87 (&net, 256, 1, 1, 1, 1, 0, 0, c87_bin, true); tk::dnn::Activation a87 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c88 (&net, 512, 3, 3, 1, 1, 1, 1, c88_bin, true); @@ -256,7 +254,6 @@ int main() { tk::dnn::Activation a90 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c91 (&net, 256, 1, 1, 1, 1, 0, 0, c91_bin, true); tk::dnn::Activation a91 (&net, tk::dnn::ACTIVATION_LEAKY); -/* tk::dnn::Conv2d c92 (&net, 512, 3, 3, 1, 1, 1, 1, c92_bin, true); tk::dnn::Activation a92 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c93 (&net, 45, 1, 1, 1, 1, 0, 0, c93_bin, false); @@ -264,14 +261,12 @@ int main() { tk::dnn::Layer *m95_layers[1] = { &a91 }; tk::dnn::Route m95 (&net, m95_layers, 1); -*/ tk::dnn::Conv2d c96 (&net, 128, 1, 1, 1, 1, 0, 0, c96_bin, true); tk::dnn::Activation a96 (&net, tk::dnn::ACTIVATION_LEAKY); - tk::dnn::Upsample u97 (&net, 2); -// tk::dnn::Layer *m98_layers[2] = { &u97, &s36 }; -// tk::dnn::Route m98 (&net, m98_layers, 2); + tk::dnn::Layer *m98_layers[2] = { &u97, &s36 }; + tk::dnn::Route m98 (&net, m98_layers, 2); tk::dnn::Conv2d c99 (&net, 128, 1, 1, 1, 1, 0, 0, c99_bin, true); tk::dnn::Activation a99 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Conv2d c100 (&net, 256, 3, 3, 1, 1, 1, 1, c100_bin, true); @@ -288,11 +283,6 @@ int main() { tk::dnn::Conv2d c105 (&net, 45, 1, 1, 1, 1, 0, 0, c105_bin, false); tk::dnn::Yolo y106 (&net, 10, 3); - // merge all yolos -// tk::dnn::Layer *m107_layers[2] = { &y82, &y94, &y106 }; -// tk::dnn::Route m107 (&net, m107_layers, 3); - - // Load input dnnType *data; dnnType *input_h;