From 10b71606772584b8367eda08d7f889273c21a4d4 Mon Sep 17 00:00:00 2001 From: Francesco Gatti Date: Sun, 16 Feb 2020 16:28:39 +0100 Subject: [PATCH] works but it need cleaning --- include/tkDNN/Layer.h | 3 +- include/tkDNN/utils.h | 2 +- src/LSTM.cpp | 163 ++++++++++++++++++++++++++++------- src/utils.cpp | 4 +- tests/imuodom/imuodom.cpp | 26 +++--- tests/imuodom/infer.py | 25 +++--- tests/simple/test_model.py | 15 ++-- tests/simple/test_simple.cpp | 8 +- tests/weights_exporter.py | 44 ++++++++-- 9 files changed, 219 insertions(+), 71 deletions(-) diff --git a/include/tkDNN/Layer.h b/include/tkDNN/Layer.h index 319df6a..f12027f 100644 --- a/include/tkDNN/Layer.h +++ b/include/tkDNN/Layer.h @@ -237,7 +237,7 @@ public: virtual dnnType* infer(dataDim_t &dim, dnnType* srcData); - const bool bidirectional = false; /**> is the net bidir */ + const bool bidirectional = true; /**> is the net bidir */ bool returnSeq = false; /**> if false return only the result of last timestep */ int stateSize = 0; /**> number of hidden states */ int seqLen = 0; /**> number of timesteps */ @@ -260,6 +260,7 @@ protected: cudnnFilterDescriptor_t w_desc_; dnnType *w_ptr; dnnType *w_h; + dnnType *wf_ptr, *wb_ptr; // params pointer forward and backward layer }; diff --git a/include/tkDNN/utils.h b/include/tkDNN/utils.h index dc34a31..3fa9d34 100644 --- a/include/tkDNN/utils.h +++ b/include/tkDNN/utils.h @@ -91,7 +91,7 @@ void printCenteredTitle(const char *title, char fill, int dim); bool fileExist(const char *fname); void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek = 0); -int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true); +int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true, int limit = 10); void printDeviceVector(int size, dnnType* vec_d, bool device = true); void resize(int size, dnnType **data); diff --git a/src/LSTM.cpp b/src/LSTM.cpp index ce2a9f5..00d8f76 100644 --- a/src/LSTM.cpp +++ b/src/LSTM.cpp @@ -37,7 +37,7 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig net->dataType, 3, dimA, strideA)); dimA[0] = batchSize; - dimA[1] = bidirectional ? stateSize*2 : stateSize; + dimA[1] = stateSize; dimA[2] = 1; strideA[0] = dimA[2] * dimA[1]; strideA[1] = dimA[2]; @@ -51,7 +51,7 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig // set the state tensors - dimA[0] = numLayers * (bidirectional ? 2 : 1); + dimA[0] = numLayers; dimA[1] = batchSize; dimA[2] = stateSize; strideA[0] = dimA[2] * dimA[1]; @@ -91,7 +91,8 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig checkCUDNN(cudnnSetRNNDescriptor(net->cudnnHandle, rnnDesc, stateSize, numLayers, dropoutDesc, cudnnRNNInputMode_t::CUDNN_LINEAR_INPUT, - (bidirectional ? cudnnDirectionMode_t::CUDNN_BIDIRECTIONAL : cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL), + //(bidirectional ? cudnnDirectionMode_t::CUDNN_BIDIRECTIONAL : cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL), + cudnnDirectionMode_t::CUDNN_UNIDIRECTIONAL, cudnnRNNMode_t::CUDNN_LSTM, cudnnRNNAlgo_t::CUDNN_RNN_ALGO_STANDARD, net->dataType)); @@ -119,23 +120,26 @@ LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weig net->dataType, net->tensorFormat, 3, dim_w)); // load params - readBinaryFile(fname_weights, cudnn_params, &w_h, &w_ptr); - - //allocate data for infer result - int dstDim = input_dim.n * stateSize*(bidirectional ? 2 : 1) * input_dim.h * input_dim.w; - checkCuda( cudaMalloc(&dstData, dstDim*sizeof(dnnType)) ); + readBinaryFile(fname_weights, cudnn_params*2, &w_h, &w_ptr); + // set forward and backward params + wf_ptr = w_ptr; + wb_ptr = w_ptr + cudnn_params; + std::cout<<"wf: "<cublasHandle, srcData, trans, dim.c, dim.h*dim.w*dim.l); + srcData = trans; + + // reposition in invered order + dnnType *srcBack; + checkCuda( cudaMalloc(&srcBack, dim.tot()*sizeof(dnnType))); + for(int i=0; icudnnHandle, + rnnDesc, + seqLen, // number of time steps (nT) + x_desc_vec_.data(), // input array of desc (nT*nC_in) + srcData, // input pointer + hx_desc_, // initial hidden state desc + hx_ptr, // initial hidden state pointer + cx_desc_, // initial cell state desc + cx_ptr, // initial cell state pointer + w_desc_, // weights desc + wf_ptr, // weights pointer + y_desc_vec_.data(), // output desc (nT*nC_out) + dstF, // output pointer + hy_desc_, // final hidden state desc + hy_ptr, // final hidden state pointer + cy_desc_, // final cell state desc + cy_ptr, // final cell state pointer + work_space_, // workspace pointer + workspace_byte_)); // workspace size + } + std::cout<<"OUTPUT F:\n"; + printDeviceVector(singleOutput.tot(), dstF); + + std::cout<<"INPUT:\n"; + printDeviceVector(input_dim.tot(), srcBack); + + // backward + { + // reset states + checkCuda( cudaMemset(hx_ptr, 0, stateDataDim*sizeof(float)) ); + checkCuda( cudaMemset(cx_ptr, 0, stateDataDim*sizeof(float)) ); + + checkCUDNN(cudnnRNNForwardInference(net->cudnnHandle, + rnnDesc, + seqLen, // number of time steps (nT) + x_desc_vec_.data(), // input array of desc (nT*nC_in) + srcBack, // input pointer + hx_desc_, // initial hidden state desc + hx_ptr, // initial hidden state pointer + cx_desc_, // initial cell state desc + cx_ptr, // initial cell state pointer + w_desc_, // weights desc + wb_ptr, // weights pointer + y_desc_vec_.data(), // output desc (nT*nC_out) + dstB, // output pointer + hy_desc_, // final hidden state desc + hy_ptr, // final hidden state pointer + cy_desc_, // final cell state desc + cy_ptr, // final cell state pointer + work_space_, // workspace pointer + workspace_byte_)); // workspace size + } + + + // reposition in invered order + dnnType *dstBack; + checkCuda( cudaMalloc(&dstBack, singleOutput.tot()*sizeof(dnnType))); + for(int i=0; icublasHandle, dstF, trans, + singleOutput.h*singleOutput.w*singleOutput.l, singleOutput.c); + // backward transpose + matrixTranspose(net->cublasHandle, dstB, trans + singleOutput.tot(), + singleOutput.h*singleOutput.w*singleOutput.l, singleOutput.c); + dstData = trans; + } else { + // copy last of forward + checkCuda( cudaMemcpy(trans, dstF + singleOutput.tot() - singleOutput.c, singleOutput.c*sizeof(dnnType), cudaMemcpyDeviceToDevice)); + // copy first of backward + checkCuda( cudaMemcpy(trans + singleOutput.c, dstB, singleOutput.c*sizeof(dnnType), cudaMemcpyDeviceToDevice)); + dstData = trans; + } - checkCUDNN(cudnnRNNForwardInference(net->cudnnHandle, - rnnDesc, - seqLen, // number of time steps (nT) - x_desc_vec_.data(), // input array of desc (nT*nC_in) - srcData, // input pointer - hx_desc_, // initial hidden state desc - hx_ptr, // initial hidden state pointer - cx_desc_, // initial cell state desc - cx_ptr, // initial cell state pointer - w_desc_, // weights desc - w_ptr, // weights pointer - y_desc_vec_.data(), // output desc (nT*nC_out) - dstData, // output pointer - hy_desc_, // final hidden state desc - hy_ptr, // final hidden state pointer - cy_desc_, // final cell state desc - cy_ptr, // final cell state pointer - work_space_, // workspace pointer - workspace_byte_)); // workspace size dim = output_dim; return dstData; diff --git a/src/utils.cpp b/src/utils.cpp index 444318d..6789e8c 100644 --- a/src/utils.cpp +++ b/src/utils.cpp @@ -68,7 +68,7 @@ void printDeviceVector(int size, dnnType* vec_d, bool device) delete [] vec; } -int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device) { +int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device, int limit) { dnnType *data_h, *correct_h; const float eps = 0.02f; @@ -92,7 +92,7 @@ int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device) { diffs += 1; if(diffs == 1) std::cout<<"\n"; - if(diffs < 10) + if(diffs < limit) std::cout<<" | [ "<