diff --git a/.gitignore b/.gitignore index 02f2a8e..68b4b70 100644 --- a/.gitignore +++ b/.gitignore @@ -8,4 +8,6 @@ build/ *.h5 *.tar.gz *.weights -.idea/ \ No newline at end of file +.idea/ +*.hdf5 +*.pk \ No newline at end of file diff --git a/CMakeLists.txt b/CMakeLists.txt index b388213..5f3052c 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -30,6 +30,9 @@ cuda_add_library(kernels SHARED ${tkdnn_CUSRC}) #------------------------------------------------------------------------------- # External Libraries #------------------------------------------------------------------------------- +find_package(Eigen3 REQUIRED) +include_directories(${EIGEN3_INCLUDE_DIR}) + find_package(OpenCV REQUIRED) set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DOPENCV") @@ -130,6 +133,8 @@ target_link_libraries(test_dla34 tkDNN) add_executable(test_dla34_cnet tests/dla34_cnet/dla34_cnet.cpp) target_link_libraries(test_dla34_cnet tkDNN) +add_executable(test_imuodom tests/imuodom/imuodom.cpp) +target_link_libraries(test_imuodom tkDNN) ################################################################################ diff --git a/include/tkDNN/ImuOdom.h b/include/tkDNN/ImuOdom.h new file mode 100644 index 0000000..a6b449c --- /dev/null +++ b/include/tkDNN/ImuOdom.h @@ -0,0 +1,143 @@ +#include +#include +#include /* srand, rand */ +#include +#include +#include +#include "utils.h" +#include "tkdnn.h" + +namespace tk { namespace dnn { + +/** + * + * @author Francesco Gatti + */ +class ImuOdom { + + public: + tk::dnn::Network *net = nullptr; + + // Network input dim + tk::dnn::dataDim_t dim0; + tk::dnn::dataDim_t dim1; + tk::dnn::dataDim_t dim2; + + // Network output dim + tk::dnn::dataDim_t odim0; + tk::dnn::dataDim_t odim1; + + // input pointers + dnnType *i0_d, *i1_d, *i2_d; + // output pointers + dnnType *o0_d, *o1_d; + + // output eigen CPU + Eigen::MatrixXf deltaP, deltaQ; + + Eigen::MatrixXd odomPOS, odomROT; + Eigen::Isometry3f tf = Eigen::Isometry3f::Identity(); + + ImuOdom() {} + + virtual ~ImuOdom() {} + + /** + * Method used for inizialize the class + * + * @return Success of the initialization + */ + bool init(std::string layers_path) { + + dim0 = tk::dnn::dataDim_t(1, 4, 1, 100); + dim1 = tk::dnn::dataDim_t(1, 3, 1, 100); + dim2 = tk::dnn::dataDim_t(1, 3, 1, 100); + + checkCuda( cudaMalloc(&i0_d, dim0.tot()*sizeof(dnnType)) ); + checkCuda( cudaMalloc(&i1_d, dim1.tot()*sizeof(dnnType)) ); + checkCuda( cudaMalloc(&i2_d, dim2.tot()*sizeof(dnnType)) ); + + std::string c0_bin = layers_path + "/conv1d_7.bin"; + std::string c1_bin = layers_path + "/conv1d_8.bin"; + std::string c2_bin = layers_path + "/conv1d_9.bin"; + std::string c3_bin = layers_path + "/conv1d_10.bin"; + std::string c4_bin = layers_path + "/conv1d_11.bin"; + std::string c5_bin = layers_path + "/conv1d_12.bin"; + std::string l0_bin = layers_path + "/bidirectional_3.bin"; + std::string l1_bin = layers_path + "/bidirectional_4.bin"; + std::string d0_bin = layers_path + "/dense_3.bin"; + std::string d1_bin = layers_path + "/dense_4.bin"; + + net = new tk::dnn::Network(dim0); + tk::dnn::Input *x0 = new tk::dnn::Input (net, dim0, i0_d); + tk::dnn::Conv2d *x0_0 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c0_bin); + tk::dnn::Conv2d *x0_1 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c1_bin); + tk::dnn::Pooling *x0_2 = new tk::dnn::Pooling(net, 1, 3, 1, 3, tk::dnn::tkdnnPoolingMode_t::POOLING_MAX); + + tk::dnn::Input *x1 = new tk::dnn::Input (net, dim1, i1_d); + tk::dnn::Conv2d *x1_0 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c2_bin); + tk::dnn::Conv2d *x1_1 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c3_bin); + tk::dnn::Pooling *x1_2 = new tk::dnn::Pooling(net, 1, 3, 1, 3, tk::dnn::tkdnnPoolingMode_t::POOLING_MAX); + + tk::dnn::Input *x2 = new tk::dnn::Input (net, dim2, i2_d); + tk::dnn::Conv2d *x2_0 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c4_bin); + tk::dnn::Conv2d *x2_1 = new tk::dnn::Conv2d (net, 128, 1, 11, 1, 1, 0, 0, c5_bin); + tk::dnn::Pooling *x2_2 = new tk::dnn::Pooling(net, 1, 3, 1, 3, tk::dnn::tkdnnPoolingMode_t::POOLING_MAX); + + tk::dnn::Layer *concat_l[3] = { x0_2, x1_2, x2_2 }; + tk::dnn::Route *concat = new tk::dnn::Route(net, concat_l, 3); + + tk::dnn::LSTM *lstm0 = new tk::dnn::LSTM(net, 128, true, l0_bin); + tk::dnn::LSTM *lstm1 = new tk::dnn::LSTM(net, 128, false, l1_bin); + + tk::dnn::Dense *d0 = new tk::dnn::Dense(net, 3, d0_bin); + + tk::dnn::Layer *lstm1_l[1] = { lstm1 }; + tk::dnn::Route *lstm1_link = new tk::dnn::Route(net, lstm1_l, 1); + tk::dnn::Dense *d1 = new tk::dnn::Dense(net, 4, d1_bin); + + net->print(); + + // output data + o0_d = d0->dstData; + o1_d = d1->dstData; + odim0 = d0->output_dim; + odim1 = d1->output_dim; + + deltaP.resize(odim0.tot(), 1); + deltaQ.resize(odim1.tot(), 1); + + odomPOS = Eigen::MatrixXd::Zero(3, 1); + odomROT = Eigen::MatrixXd::Identity(3, 3); + } + + void update(dnnType *x0, dnnType *x1, dnnType *x2) { + + checkCuda( cudaMemcpy(i0_d, x0, dim0.tot()*sizeof(dnnType), cudaMemcpyHostToDevice) ); + checkCuda( cudaMemcpy(i1_d, x1, dim1.tot()*sizeof(dnnType), cudaMemcpyHostToDevice) ); + checkCuda( cudaMemcpy(i2_d, x2, dim2.tot()*sizeof(dnnType), cudaMemcpyHostToDevice) ); + + // Inference + tk::dnn::dataDim_t dim; + net->infer(dim, nullptr); + + checkCuda( cudaMemcpy(deltaP.data(), o0_d, odim0.tot()*sizeof(dnnType), cudaMemcpyDeviceToHost) ); + checkCuda( cudaMemcpy(deltaQ.data(), o1_d, odim1.tot()*sizeof(dnnType), cudaMemcpyDeviceToHost) ); + + // compute odom + Eigen::Quaterniond q; + q.w() = deltaQ(0); + q.x() = deltaQ(1); + q.y() = deltaQ(2); + q.z() = deltaQ(3); + odomPOS = odomPOS + odomROT*deltaP.cast(); + odomROT = odomROT * q.normalized().toRotationMatrix(); + + // compose tf + tf.matrix().block(0, 0, 3, 3) = odomROT.cast(); + tf.matrix().block(0, 3, 3, 1) = odomPOS.cast(); + } + +}; + +}} diff --git a/include/tkDNN/Layer.h b/include/tkDNN/Layer.h index 910435b..c1167af 100644 --- a/include/tkDNN/Layer.h +++ b/include/tkDNN/Layer.h @@ -9,10 +9,12 @@ namespace tk { namespace dnn { enum layerType_t { + LAYER_INPUT, LAYER_DENSE, LAYER_CONV2D, LAYER_DECONV2D, LAYER_DEFORMCONV2D, + LAYER_LSTM, LAYER_ACTIVATION, LAYER_ACTIVATION_CRELU, LAYER_ACTIVATION_LEAKY, @@ -55,10 +57,12 @@ public: std::string getLayerName() { layerType_t type = getLayerType(); switch(type) { + case LAYER_INPUT: return "Input"; case LAYER_DENSE: return "Dense"; case LAYER_CONV2D: return "Conv2d"; case LAYER_DECONV2D: return "DeConv2d"; case LAYER_DEFORMCONV2D: return "DeformConv2d"; + case LAYER_LSTM: return "LSTM"; case LAYER_ACTIVATION: return "Activation"; case LAYER_ACTIVATION_CRELU: return "ActivationCReLU"; case LAYER_ACTIVATION_LEAKY: return "ActivationLeaky"; @@ -123,6 +127,28 @@ public: }; +/** + Input layer (it doesnt need weigths) +*/ +class Input : public Layer { + +public: + + Input(Network *net, dataDim_t &dim, dnnType* srcData) : Layer(net) { + input_dim = dim; + output_dim = dim; + dstData = srcData; + } + virtual ~Input() {} + virtual layerType_t getLayerType() { return LAYER_INPUT; }; + + virtual dnnType* infer(dataDim_t &dim, dnnType* srcData) { + dim = output_dim; + return dstData; + } +}; + + /** Dense (full interconnection) layer */ @@ -174,6 +200,14 @@ protected: /** Convolutional 2D layer + + WEIGHTS shape: OUTCH, INCH, KH, KW ... + BIAS shape: OUTCH + + with BATCHNORM: + scales: OUTCH + means: OUTCH + variance: OUTCH */ class Conv2d : public LayerWgs { @@ -203,6 +237,71 @@ protected: size_t ws_sizeInBytes; }; +/** + Bidirectional LSTM layer + ONLY BIDIRECTIONAL (TODO: more configurable) + currently implemented as 2 inferences: forward and backward (TODO: only 1 cudnn inference) + + implementation info: + https://github.com/jiangnanhugo/seq2seq_cuda/blob/e4dbdcfa0517c972bfd4beea9f11a5233954093c/src/rnn.cpp + https://github.com/Jeffery-Song/mxnet-test/blob/aab666faad44011f7a67b527b5f6c960367d0422/src/operator/cudnn_rnn-inl.h + https://stackoverflow.com/a/38737941 + https://colah.github.io/posts/2015-08-Understanding-LSTMs/ + + PARAMS (numlayers*2): + layer0: + ( INCH, ? ) ??? + ( HIDDEN, ? ) ??? + ( HIDDEN * 8 ) ??? + layer2: + ( INCH, ? ) ??? + ( HIDDEN, ? ) ??? + ( HIDDEN * 8 ) ??? + + OUTPUT shape: + (N, C, 1, W) ---> LSTM(HIDDEN, returnSeq=True) ---> (N, 2*HIDDEN, 1, W) # W is seqLength + (N, C, 1, W) ---> LSTM(HIDDEN, returnSeq=False) ---> (N, 2*HIDDEN, 1, 1) +*/ +class LSTM : public Layer { + +public: + LSTM(Network *net, int hiddensize, bool returnSeq, std::string fname_weights); + virtual ~LSTM(); + virtual layerType_t getLayerType() { return LAYER_LSTM; }; + + virtual dnnType* infer(dataDim_t &dim, dnnType* srcData); + + 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 */ + int numLayers = 1; /**> number of internal layers */ + +protected: + cudnnRNNDescriptor_t rnnDesc; + cudnnDropoutDescriptor_t dropoutDesc; + dnnType *dropout_states_, *work_space_; + + size_t workspace_byte_, dropout_byte_; + int workspace_size_, dropout_size_; + + std::vector x_desc_vec_, y_desc_vec_; + cudnnTensorDescriptor_t hx_desc_, cx_desc_; + cudnnTensorDescriptor_t hy_desc_, cy_desc_; + dnnType *hx_ptr, *cx_ptr, *hy_ptr, *cy_ptr; + int stateDataDim; + + cudnnFilterDescriptor_t w_desc_; + dnnType *w_ptr; + dnnType *w_h; + dnnType *wf_ptr, *wb_ptr; // params pointer forward and backward layer + + // used during inference + dataDim_t one_output_dim; // output dim of as single inference + dnnType *srcF, *srcB; // input of single inference + dnnType *dstF, *dstB_NR, *dstB; // output of single inference, dstB_NR = dstB not reversed +}; + /** Convolutional 2D layer @@ -370,8 +469,8 @@ public: virtual dnnType* infer(dataDim_t &dim, dnnType* srcData); public: - static const int MAX_INPUT_LAYERS = 16; - Layer *layers[MAX_INPUT_LAYERS]; //ids of layers to be merged + static const int MAX_LAYERS = 32; + Layer *layers[MAX_LAYERS]; //ids of layers to be merged int layers_n; //number of layers }; diff --git a/include/tkDNN/utils.h b/include/tkDNN/utils.h index a205870..0fed5c9 100644 --- a/include/tkDNN/utils.h +++ b/include/tkDNN/utils.h @@ -93,11 +93,11 @@ } \ } -void printCenteredTitle(const char *title, char fill, int dim); +void printCenteredTitle(const char *title, char fill, int dim = 30); bool fileExist(const char *fname); void downloadWeightsifDoNotExist(const std::string& input_bin, const std::string& test_folder, const std::string& weights_url); -void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek = 0, bool skipLoad = false); -int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true); +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 limit = 10); void printDeviceVector(int size, dnnType* vec_d, bool device = true); float getColor(const int c, const int x, const int max); void resize(int size, dnnType **data); diff --git a/scripts/test_all_tests.sh b/scripts/test_all_tests.sh index a51fdd5..a4bb539 100644 --- a/scripts/test_all_tests.sh +++ b/scripts/test_all_tests.sh @@ -1,6 +1,7 @@ #!/bin/bash cd build +# rm *rt RED='\033[1;31m' GREEN='\033[1;32m' @@ -31,9 +32,9 @@ print_output $res_yolo3 test_yolo3 res_yolo3_flir=$? print_output $res_yolo3_flir test_yolo3_flir -./test_yolo3_512 &>> $out_file -res_yolo3_512=$? -print_output $res_yolo3_512 test_yolo3_512 +# ./test_yolo3_512 &>> $out_file +# res_yolo3_512=$? +# print_output $res_yolo3_512 test_yolo3_512 ./test_yolo3_tiny &>> $out_file res_yolo3_tiny=$? @@ -67,20 +68,20 @@ print_output $res_mnist test_mnist res_yolo=$? print_output $res_yolo test_yolo -./test_yolo3_berkeley &>> $out_file -res_yolo3_berkeley=$? -print_output $res_yolo3_berkeley test_yolo3_berkeley +# ./test_yolo3_berkeley &>> $out_file +# res_yolo3_berkeley=$? +# print_output $res_yolo3_berkeley test_yolo3_berkeley ./test_yolo_voc &>> $out_file res_yolo_voc=$? print_output $res_yolo_voc test_yolo_voc -./test_dla34_cnet &>> $out_file -res_dla34_cnet=$? -print_output $res_dla34_cnet test_dla34_cnet +# ./test_dla34_cnet &>> $out_file +# res_dla34_cnet=$? +# print_output $res_dla34_cnet test_dla34_cnet -./test_yolo3_coco4 &>> $out_file -res_yolo3_coco4=$? -print_output $res_yolo3_coco4 test_yolo3_coco4 +# ./test_yolo3_coco4 &>> $out_file +# res_yolo3_coco4=$? +# print_output $res_yolo3_coco4 test_yolo3_coco4 echo "If errors occured, check logfile $out_file" diff --git a/src/LSTM.cpp b/src/LSTM.cpp new file mode 100644 index 0000000..6ecf4fd --- /dev/null +++ b/src/LSTM.cpp @@ -0,0 +1,327 @@ +#include + +#include "Layer.h" + +namespace tk { namespace dnn { + +LSTM::LSTM( Network *net, int hiddensize, bool returnSeq, std::string fname_weights) : + Layer(net) { + + this->returnSeq = returnSeq; + int batchSize = input_dim.n; + int inputSize = input_dim.c; + seqLen = input_dim.w; + stateSize = hiddensize; + + // init Tensor Descriptors + std::vector x_vec(seqLen); + std::vector y_vec(seqLen); + + int dimA[3]; + int strideA[3]; + for (int i = 0; i < seqLen; i++) { + checkCUDNN(cudnnCreateTensorDescriptor(&x_vec[i])); + checkCUDNN(cudnnCreateTensorDescriptor(&y_vec[i])); + + dimA[0] = batchSize; + dimA[1] = inputSize; + dimA[2] = 1; + dimA[0] = batchSize; + dimA[1] = inputSize; + strideA[0] = dimA[2] * dimA[1]; + strideA[1] = dimA[2]; + strideA[2] = 1; + checkCUDNN(cudnnSetTensorNdDescriptor(x_vec[i], + net->dataType, 3, dimA, strideA)); + + dimA[0] = batchSize; + dimA[1] = stateSize; + dimA[2] = 1; + strideA[0] = dimA[2] * dimA[1]; + strideA[1] = dimA[2]; + strideA[2] = 1; + checkCUDNN(cudnnSetTensorNdDescriptor(y_vec[i], + net->dataType, 3, dimA, strideA)); + } + // apply tensordesc + x_desc_vec_ = x_vec; + y_desc_vec_ = y_vec; + + + // set the state tensors + dimA[0] = numLayers; + dimA[1] = batchSize; + dimA[2] = stateSize; + strideA[0] = dimA[2] * dimA[1]; + strideA[1] = dimA[2]; + strideA[2] = 1; + checkCUDNN(cudnnCreateTensorDescriptor(&hx_desc_)); + checkCUDNN(cudnnCreateTensorDescriptor(&cx_desc_)); + checkCUDNN(cudnnCreateTensorDescriptor(&hy_desc_)); + checkCUDNN(cudnnCreateTensorDescriptor(&cy_desc_)); + checkCUDNN(cudnnSetTensorNdDescriptor(hx_desc_, net->dataType, 3, dimA, strideA)); + checkCUDNN(cudnnSetTensorNdDescriptor(cx_desc_, net->dataType, 3, dimA, strideA)); + checkCUDNN(cudnnSetTensorNdDescriptor(hy_desc_, net->dataType, 3, dimA, strideA)); + checkCUDNN(cudnnSetTensorNdDescriptor(cy_desc_, net->dataType, 3, dimA, strideA)); + // allocate dnnType *hx_ptr, *cx_ptr, *hy_ptr, *cy_ptr; + stateDataDim = dimA[0]*dimA[1]*dimA[2]; + checkCuda( cudaMalloc(&hx_ptr, stateDataDim*sizeof(dnnType)) ); + checkCuda( cudaMalloc(&cx_ptr, stateDataDim*sizeof(dnnType)) ); + checkCuda( cudaMalloc(&hy_ptr, stateDataDim*sizeof(dnnType)) ); + checkCuda( cudaMalloc(&cy_ptr, stateDataDim*sizeof(dnnType)) ); + + + + // Create Dropout descriptors // TODO: ??? IS IT NECESSARY ??? + float dropoutprob = 0.1f; // random val ???? + checkCUDNN(cudnnCreateDropoutDescriptor(&dropoutDesc)); + checkCUDNN(cudnnDropoutGetStatesSize(net->cudnnHandle, &dropout_byte_)); + dropout_size_ = dropout_byte_ / sizeof(dnnType); + checkCuda( cudaMalloc(&dropout_states_, dropout_byte_) ); + uint64_t seed_ = 17 + rand() % 4096; // NOLINT(runtime/threadsafe_fn) + checkCUDNN(cudnnSetDropoutDescriptor(dropoutDesc, + net->cudnnHandle, dropoutprob, dropout_states_, dropout_byte_, seed_)); + + + // RNN descriptors + checkCUDNN(cudnnCreateRNNDescriptor(&rnnDesc)); + + checkCUDNN(cudnnSetRNNDescriptor(net->cudnnHandle, + rnnDesc, stateSize, numLayers, dropoutDesc, + cudnnRNNInputMode_t::CUDNN_LINEAR_INPUT, + //(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)); + + + // Get temp space sizes + checkCUDNN(cudnnGetRNNWorkspaceSize(net->cudnnHandle, + rnnDesc, seqLen, x_desc_vec_.data(), &workspace_byte_)); + workspace_size_ = workspace_byte_ / sizeof(dnnType); + checkCuda( cudaMalloc(&work_space_, workspace_byte_) ); + + + // Check that number of params are correct + size_t cudnn_param_size; + checkCUDNN(cudnnGetRNNParamsSize(net->cudnnHandle, + rnnDesc,x_desc_vec_[0], &cudnn_param_size, net->dataType)); + int cudnn_params = cudnn_param_size/sizeof(dnnType); + //std::cout<<"LSTM params size: "<dataType, net->tensorFormat, 3, dim_w)); + + // load params + std::cout<<"Reading weights: PARAMS="<cudnnHandle, rnnDesc, + i, x_desc_vec_[0], w_desc_, 0, j, m_desc, (void**)&p)); + + std::cout << "ptr: " << ((int64_t)(p - NULL))/sizeof(dnnType)<<"\n"; + + cudnnDataType_t t; + cudnnTensorFormat_t f; + int ndim = 5; + int dims[5] = {0, 0, 0, 0, 0}; + checkCUDNN(cudnnGetFilterNdDescriptor(m_desc, ndim, &t, &f, &ndim, &dims[0])); + std::cout << "(layer, linlayer): " << i << " " << j << "\n"; + + int tot = 1; + for (int i = 0; i < ndim; ++i) { + std::cout << dims[i] << " "; + tot *= dims[i]; + } + std::cout<<"\t-> "<cublasHandle, srcData, srcF, dim.c, dim.h*dim.w*dim.l); + + // build srcB as reversed srcF + for(int i=0; icudnnHandle, + rnnDesc, + seqLen, // number of time steps (nT) + x_desc_vec_.data(), // input array of desc (nT*nC_in) + srcF, // 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 + } + + // 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) + srcB, // 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_NR, // 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 + } + + + // reverse order of dstB + for(int i=0; icublasHandle, dstF, dstData, + one_output_dim.h* one_output_dim.w*one_output_dim.l, one_output_dim.c); + // backward transpose + matrixTranspose(net->cublasHandle, dstB, dstData + one_output_dim.tot(), + one_output_dim.h* one_output_dim.w*one_output_dim.l, one_output_dim.c); + } else { + // copy last of forward + checkCuda( cudaMemcpy(dstData, dstF + one_output_dim.tot() - one_output_dim.c, + one_output_dim.c*sizeof(dnnType), cudaMemcpyDeviceToDevice)); + // copy first of backward + checkCuda( cudaMemcpy(dstData + one_output_dim.c, dstB, + one_output_dim.c*sizeof(dnnType), cudaMemcpyDeviceToDevice)); + } + + dim = output_dim; + return dstData; +} + +}} diff --git a/src/LayerWgs.cpp b/src/LayerWgs.cpp index ed1b62c..59bf126 100644 --- a/src/LayerWgs.cpp +++ b/src/LayerWgs.cpp @@ -17,24 +17,24 @@ LayerWgs::LayerWgs(Network *net, int inputs, int outputs, std::cout<<"Reading weights: I="<dontLoadWeights); + readBinaryFile(weights_path.c_str(), inputs*outputs*kh*kw*kl, &data_h, &data_d, seek); seek += inputs*outputs*kh*kw*kl; this->additional_bias = additional_bias; if(additional_bias) { - readBinaryFile(weights_path.c_str(), outputs, &bias2_h, &bias2_d, seek, net->dontLoadWeights); + readBinaryFile(weights_path.c_str(), outputs, &bias2_h, &bias2_d, seek); seek += outputs; } - readBinaryFile(weights_path.c_str(), outputs, &bias_h, &bias_d, seek, net->dontLoadWeights); + readBinaryFile(weights_path.c_str(), outputs, &bias_h, &bias_d, seek); this->batchnorm = batchnorm; if(batchnorm) { seek += outputs; - readBinaryFile(weights_path.c_str(), outputs, &scales_h, &scales_d, seek, net->dontLoadWeights); + readBinaryFile(weights_path.c_str(), outputs, &scales_h, &scales_d, seek); seek += outputs; - readBinaryFile(weights_path.c_str(), outputs, &mean_h, &mean_d, seek, net->dontLoadWeights); + readBinaryFile(weights_path.c_str(), outputs, &mean_h, &mean_d, seek); seek += outputs; - readBinaryFile(weights_path.c_str(), outputs, &variance_h, &variance_d, seek, net->dontLoadWeights); + readBinaryFile(weights_path.c_str(), outputs, &variance_h, &variance_d, seek); float eps = TKDNN_BN_MIN_EPSILON; diff --git a/src/Route.cpp b/src/Route.cpp index 115c39a..fb8e67a 100644 --- a/src/Route.cpp +++ b/src/Route.cpp @@ -7,11 +7,14 @@ namespace tk { namespace dnn { Route::Route(Network *net, Layer **layers, int layers_n, bool final) : Layer(net, final) { - this->layers_n = layers_n; - if(layers_n > MAX_INPUT_LAYERS) - FatalError("Route: MAX INPUT LAYERS overload"); - for(int i=0; i MAX_LAYERS) { + FatalError("ROUTE: reached max number of input layers"); + } + for(int i=0; ilayers[i] = layers[i]; + } + this->layers_n = layers_n; //get dims output_dim.l = 1; diff --git a/src/utils.cpp b/src/utils.cpp index edcb244..b196d1c 100644 --- a/src/utils.cpp +++ b/src/utils.cpp @@ -32,43 +32,41 @@ void downloadWeightsifDoNotExist(const std::string& input_bin, const std::string } -void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek, bool skipLoad){ +void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek) +{ + std::ifstream dataFile (fname, std::ios::in | std::ios::binary); + std::stringstream error_s; + if (!dataFile) + { + error_s << "Error opening file " << fname; + FatalError(error_s.str()); + } + + if(seek != 0) { + dataFile.seekg(seek*sizeof(dnnType), dataFile.cur); + } + int size_b = size*sizeof(dnnType); *data_h = new dnnType[size]; - - if(!skipLoad) { - std::ifstream dataFile(fname, std::ios::in | std::ios::binary); - std::stringstream error_s; - if (!dataFile) { - error_s << "Error opening file " << fname; - FatalError(error_s.str()); - } - - if (seek != 0) { - dataFile.seekg(seek * sizeof(dnnType), dataFile.cur); - } - - // printf("data_h %d size_b %d\n", *data_h,size_b); - if (!dataFile.read((char *) *data_h, size_b)) { - - error_s << "Error reading file " << fname; - FatalError(error_s.str()); - } - } else { - std::cout< +#include "tkDNN/ImuOdom.h" + +const char *i0_bin = "../tests/imuodom/layers/input0.bin"; +const char *i1_bin = "../tests/imuodom/layers/input1.bin"; +const char *i2_bin = "../tests/imuodom/layers/input2.bin"; +const char *o0_bin = "../tests/imuodom/layers/output0.bin"; +const char *o1_bin = "../tests/imuodom/layers/output1.bin"; + +const char *c0_bin = "../tests/imuodom/layers/conv1d_7.bin"; +const char *c1_bin = "../tests/imuodom/layers/conv1d_8.bin"; +const char *c2_bin = "../tests/imuodom/layers/conv1d_9.bin"; +const char *c3_bin = "../tests/imuodom/layers/conv1d_10.bin"; +const char *c4_bin = "../tests/imuodom/layers/conv1d_11.bin"; +const char *c5_bin = "../tests/imuodom/layers/conv1d_12.bin"; +const char *l0_bin = "../tests/imuodom/layers/bidirectional_3.bin"; +const char *l1_bin = "../tests/imuodom/layers/bidirectional_4.bin"; +const char *d0_bin = "../tests/imuodom/layers/dense_3.bin"; +const char *d1_bin = "../tests/imuodom/layers/dense_4.bin"; + + +int main() { + + tk::dnn::ImuOdom ImuNet; + ImuNet.init("../tests/imuodom/layers/"); + + const int N = 19513; + + // Network layout + tk::dnn::dataDim_t dim0(1, 4, 1, 100); + tk::dnn::dataDim_t dim1(1, 3, 1, 100); + tk::dnn::dataDim_t dim2(1, 3, 1, 100); + + // Load input + dnnType *i0_d, *i1_d, *i2_d; + dnnType *i0_h, *i1_h, *i2_h; + readBinaryFile(i0_bin, dim0.tot()*N, &i0_h, &i0_d); + readBinaryFile(i1_bin, dim1.tot()*N, &i1_h, &i1_d); + readBinaryFile(i2_bin, dim2.tot()*N, &i2_h, &i2_d); + + dnnType *data; + tk::dnn::dataDim_t dim; + + dnnType *out0, *out1; + dnnType *out0_h, *out1_h; + readBinaryFile(o0_bin, ImuNet.odim0.tot()*N, &out0_h, &out0); + readBinaryFile(o1_bin, ImuNet.odim1.tot()*N, &out1_h, &out1); + + + std::ofstream path("path.txt"); + + for(int i=0; i