From ed029304648a92b9fd55cd92f8d24ad4ddac7c8d Mon Sep 17 00:00:00 2001 From: Francesco Gatti Date: Wed, 19 Dec 2018 22:45:43 +0100 Subject: [PATCH] upsample template --- include/Layer.h | 17 +++++++++++++++++ src/Upsample.cpp | 33 +++++++++++++++++++++++++++++++++ 2 files changed, 50 insertions(+) create mode 100644 src/Upsample.cpp diff --git a/include/Layer.h b/include/Layer.h index d4d0378..6d59a03 100644 --- a/include/Layer.h +++ b/include/Layer.h @@ -18,6 +18,7 @@ enum layerType_t { LAYER_ROUTE, LAYER_REORG, LAYER_SHORTCUT, + LAYER_UPSAMPLE, LAYER_REGION, }; @@ -52,6 +53,7 @@ public: case LAYER_ROUTE: return "Route"; case LAYER_REORG: return "Reorg"; case LAYER_SHORTCUT: return "Shortcut"; + case LAYER_UPSAMPLE: return "Upsample"; case LAYER_REGION: return "Region"; default: return "unknown"; } @@ -301,6 +303,21 @@ public: Layer *backLayer; }; +/** + Upsample layer + Mantain same dimension but change C*H*W distribution +*/ +class Upsample : public Layer { + +public: + Upsample(Network *net, int stride); + virtual ~Upsample(); + virtual layerType_t getLayerType() { return LAYER_UPSAMPLE; }; + + virtual dnnType* infer(dataDim_t &dim, dnnType* srcData); + + int stride; +}; struct box { int cl; diff --git a/src/Upsample.cpp b/src/Upsample.cpp new file mode 100644 index 0000000..cb9ffbb --- /dev/null +++ b/src/Upsample.cpp @@ -0,0 +1,33 @@ +#include + +#include "Layer.h" +#include "kernels.h" + +namespace tk { namespace dnn { + +Upsample::Upsample(Network *net, int stride) : Layer(net) { + + this->stride = stride; + + output_dim.n = input_dim.n; + output_dim.c = input_dim.c*stride*stride; + output_dim.h = input_dim.h/stride; + output_dim.w = input_dim.w/stride; + output_dim.l = input_dim.l; + + checkCuda( cudaMalloc(&dstData, input_dim.tot()*sizeof(dnnType)) ); +} + +Upsample::~Upsample() { + + checkCuda( cudaFree(dstData) ); +} + +dnnType* Upsample::infer(dataDim_t &dim, dnnType* srcData) { + + + dim = output_dim; + return dstData; +} + +}}