mmr/tkDNN
Archived
1
0
Fork 0
Code Issues 46 Pull Requests 12 Releases 4 Activity

Merge with master

Browse Source 
Signed-off-by: Micaela Verucchi <micaelaverucchi@gmail.com>
This commit is contained in:
Micaela Verucchi
2020-02-27 17:41:26 +01:00
parent 6911752e3e 443691414a
commit fe37bcdf1e
41 changed files with 2179 additions and 851 deletions
Download Patch File Download Diff File Expand all files Collapse all files
.gitignore
+2
View File
@@ -15,3 +15,5 @@ build/
*.tar.gz *.tar.gz
*.weights *.weights
*.zip *.zip
.idea/
*.hdf5
CMakeLists.txt
+60 -36
View File
@@ -1,50 +1,51 @@
cmake_minimum_required(VERSION 2.8) cmake_minimum_required(VERSION 3.5)
project (tkDNN) project (tkDNN)
set(CMAKE_MODULE_PATH ${CMAKE_MODULE_PATH} ${CMAKE_CURRENT_SOURCE_DIR}/cmake)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -std=c++11 -fPIC")
include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include/tkDNN)
set(BUILD_DEPS true CACHE BOOL "If true download deps") # project specific flags
if( ${BUILD_DEPS} )
message("Launching pre-build dependency installer script...")
execute_process (COMMAND bash -c "bash build_models.sh download"
WORKING_DIRECTORY ${CMAKE_CURRENT_SOURCE_DIR}/tests)
set(BUILD_DEPS false CACHE BOOL "If true download deps" FORCE)
message("Finished dowloading test weights")
endif()
if(DEBUG) if(DEBUG)
add_definitions(-DDEBUG) add_definitions(-DDEBUG)
endif() endif()
find_package(CUDA REQUIRED)
#-------------------------------------------------------------------------------
# CUDA
#-------------------------------------------------------------------------------
find_package(CUDA 9.0 REQUIRED)
SET(CUDA_SEPARABLE_COMPILATION ON)
#set(CUDA_NVCC_FLAGS "${CUDA_NVCC_FLAGS} -arch=sm_30 --compiler-options '-fPIC'")
find_package(CUDNN REQUIRED)
# compile
file(GLOB tkdnn_CUSRC "src/kernels/*.cu")
cuda_include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS} ${CUDNN_INCLUDE_DIRS})
cuda_add_library(kernels SHARED ${tkdnn_CUSRC})
#-------------------------------------------------------------------------------
# External Libraries
#-------------------------------------------------------------------------------
find_package(OpenCV REQUIRED) find_package(OpenCV REQUIRED)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DOPENCV") set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DOPENCV")
include_directories(/usr/include/gdal) include_directories(/usr/include/gdal)
# compile Discovery only if TensorRT is installed #-------------------------------------------------------------------------------
find_library(NVINFER NAMES nvinfer) # Build Libraries
if(NVINFER STREQUAL "NVINFER-NOTFOUND") #-------------------------------------------------------------------------------
set(NVINFER_INCLUDES "/usr/local/nvidia/tensorrt/include/")
link_directories(/usr/local/nvidia/tensorrt/targets/x86_64-linux-gnu/lib/
/usr/local/cuda/targets/x86_64-linux/lib/)
endif()
file(GLOB tkdnn_CUSRC "src/kernels/*.cu")
cuda_include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS} ${NVINFER_INCLUDES})
cuda_add_library(kernels SHARED ${tkdnn_CUSRC})
file(GLOB tkdnn_SRC "src/*.cpp") file(GLOB tkdnn_SRC "src/*.cpp")
set(tkdnn_LIBS kernels ${CUDA_LIBRARIES} ${CUDA_CUBLAS_LIBRARIES} -lcudnn -lnvinfer ${OpenCV_LIBS} -lgdal) set(tkdnn_LIBS kernels ${CUDA_LIBRARIES} ${CUDA_CUBLAS_LIBRARIES} ${CUDNN_LIBRARIES} ${OpenCV_LIBS})
file(GLOB class_SRC "src/class_src/*.cpp") file(GLOB class_SRC "src/class_src/*.cpp")
set(class_LIBS ${OpenCV_LIBS} -lgdal yaml-cpp python2.7)
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -Wall -std=c++11 -O3") set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -Wall -std=c++11 -O3")
include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS} ${OPENCV_INCLUDE_DIRS} ${NVINFER_INCLUDES} "~/repos/cereal/include" ${CMAKE_CURRENT_SOURCE_DIR}/tracker_CLASS/c++/src) include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS} ${OPENCV_INCLUDE_DIRS} ${NVINFER_INCLUDES} "~/repos/cereal/include" ${CMAKE_CURRENT_SOURCE_DIR}/tracker_CLASS/c++/src /usr/include/python2.7)
include_directories( BEFORE ${MY_SOURCE_DIR}/src /usr/include/python2.7 )
set(class_LIBS ${OpenCV_LIBS} -lgdal yaml-cpp python2.7)
add_library(tkDNN SHARED ${tkdnn_SRC}) add_library(tkDNN SHARED ${tkdnn_SRC})
target_link_libraries(tkDNN ${tkdnn_LIBS}) target_link_libraries(tkDNN ${tkdnn_LIBS})
@@ -100,6 +101,11 @@ target_link_libraries(test_yolo3_tetrapack_resize tkDNN)
add_executable(test_yolo3_BCDS6 tests/yolo3_BCDS6/yolo3_BCDS6.cpp) add_executable(test_yolo3_BCDS6 tests/yolo3_BCDS6/yolo3_BCDS6.cpp)
target_link_libraries(test_yolo3_BCDS6 tkDNN) target_link_libraries(test_yolo3_BCDS6 tkDNN)
add_executable(test_yolo3_flir tests/yolo3_flir/yolo3_flir.cpp)
target_link_libraries(test_yolo3_flir tkDNN)
add_executable(test_imuodom tests/imuodom/imuodom.cpp)
target_link_libraries(test_imuodom tkDNN)
################################################################################ ################################################################################
@@ -117,15 +123,33 @@ target_link_libraries(yolo3_demo tkDNN CLASS)
#install
#-------------------------------------------------------------------------------
# Install
#-------------------------------------------------------------------------------
#if (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT) #if (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT)
# set (CMAKE_INSTALL_PREFIX "${CMAKE_BINARY_DIR}/install" # set (CMAKE_INSTALL_PREFIX "${CMAKE_BINARY_DIR}/install"
# CACHE PATH "default install path" FORCE) # CACHE PATH "default install path" FORCE)
#endif() #endif()
message("install dir:" ${CMAKE_INSTALL_PREFIX}) message("install dir:" ${CMAKE_INSTALL_PREFIX})
install(DIRECTORY include/ DESTINATION include/${CMAKE_PROJECT_NAME} install(DIRECTORY include/ DESTINATION include/)
FILES_MATCHING PATTERN "*.h")
install(TARGETS tkDNN kernels DESTINATION lib) install(TARGETS tkDNN kernels DESTINATION lib)
install(FILES "${CMAKE_SOURCE_DIR}/${CMAKE_PROJECT_NAME}Config.cmake" # source directory install(DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/cmake/" # source directory
DESTINATION "share/${CMAKE_PROJECT_NAME}/cmake/" # target directory DESTINATION "share/tkDNN/cmake/" # target directory
) )
#-------------------------------------------------------------------------------
# Prepare for test
#-------------------------------------------------------------------------------
set(TEST_DATA true CACHE BOOL "If true download deps")
if( ${TEST_DATA} )
message("Launching pre-build dependency installer script...")
execute_process (COMMAND bash -c "bash build_models.sh download"
WORKING_DIRECTORY ${CMAKE_CURRENT_SOURCE_DIR}/tests)
set(TEST_DATA false CACHE BOOL "If true download deps" FORCE)
message("Finished dowloading test weights")
endif()
README.md
+3 -3
View File
@@ -11,7 +11,7 @@ this branch actually work on every NVIDIA GPU that support the dependencies:
## Dependencies ## Dependencies
``` ```
sudo apt install libgdal-dev libeigen3-dev python-matplotlib libyaml-cpp-dev libcereal-dev sudo apt install libgdal-dev libeigen3-dev python-matplotlib libyaml-cpp-dev libcereal-dev python2.7-dev
``` ```
## Workflow ## Workflow
@@ -27,6 +27,7 @@ Build with cmake
mkdir build mkdir build
cd build cd build
cmake .. cmake ..
# use -DTEST_DATA=False to skip dataset download
make make
``` ```
during the cmake configuration it will be dowloaded the weights needed for running during the cmake configuration it will be dowloaded the weights needed for running
@@ -54,5 +55,4 @@ this will genereate a yolo3_berkeley.rt file that can be used for live detection
./yolo3_demo # launch detection on a demo video ./yolo3_demo # launch detection on a demo video
./yolo3_demo yolo3_berkeley.rt /dev/video0 # launch detection on device 0 ./yolo3_demo yolo3_berkeley.rt /dev/video0 # launch detection on device 0
``` ```
![demo](https://user-images.githubusercontent.com/11562617/72547657-540e7800-388d-11ea-83c6-49dfea2a0607.gif)
cmake/FindCUDNN.cmake
+33
View File
@@ -0,0 +1,33 @@
# Find the header files
find_path(CUDNN_INCLUDE_DIR
${CMAKE_SYSROOT}/usr/local/include
${CMAKE_SYSROOT}/usr/include
/usr/local/nvidia/tensorrt/include/
NO_DEFAULT_PATH
)
set(OLD_ROOT ${CMAKE_FIND_ROOT_PATH})
list(APPEND CMAKE_FIND_ROOT_PATH /)
list(APPEND CMAKE_FIND_LIBRARY_SUFFIXES .so.7)
list(APPEND CMAKE_FIND_LIBRARY_SUFFIXES .so.5)
find_library(CUDNN_LIB
NAMES cudnn
PATHS
/usr/local/driveworks/targets/${CMAKE_SYSTEM_PROCESSOR}-Linux/lib
/usr/lib/${CMAKE_SYSTEM_PROCESSOR}-linux-gnu/
NO_DEFAULT_PATH
)
find_library(CUDNN_NVLIB
NAMES "nvinfer"
PATHS
/usr/local/driveworks/targets/${CMAKE_SYSTEM_PROCESSOR}-Linux/lib
/usr/lib/${CMAKE_SYSTEM_PROCESSOR}-linux-gnu/
NO_DEFAULT_PATH
)
set(CMAKE_FIND_ROOT_PATH ${OLD_ROOT})
set(CUDNN_LIBRARIES ${CUDNN_LIB} ${CUDNN_NVLIB})
message("-- Found CUDNN: " ${CUDNN_LIB})
message("-- Found NVINFER: " ${CUDNN_NVLIB})
set(CUDNN_FOUND true)
cmake/tkDNNConfig.cmake
+24
View File
@@ -0,0 +1,24 @@
message("-- Found tkDNN")
set(CMAKE_MODULE_PATH ${CMAKE_CURRENT_LIST_DIR})
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} --std=c++11 -fPIC")
find_package(CUDA REQUIRED)
find_package(OpenCV REQUIRED)
find_package(CUDNN REQUIRED)
set(tkDNN_INCLUDE_DIRS
${CUDA_INCLUDE_DIRS}
${OPENCV_INCLUDE_DIRS}
${CUDNN_INCLUDE_DIRS}
)
set(tkDNN_LIBRARIES
tkDNN
kernels
${CUDA_LIBRARIES}
${CUDA_CUBLAS_LIBRARIES}
${CUDNN_LIBRARIES}
${OpenCV_LIBS}
)
set(tkDNN_FOUND true)
demo/demo/demo.cpp
+12
View File
@@ -30,6 +30,7 @@ std::string obj_class[10]{"person", "car", "truck", "bus", "motor", "bike", "rid
//mutex for some opencv operations //mutex for some opencv operations
std::mutex mutex_cv; std::mutex mutex_cv;
Show_t updates; Show_t updates;
bool SAVE_RESULT = false;
void sig_handler(int signo) void sig_handler(int signo)
{ {
@@ -302,6 +303,14 @@ void *computationTask(void *x_void_ptr)
// float prob; // float prob;
cv::Scalar intensity; cv::Scalar intensity;
// cv::VideoWriter resultVideo;
// if(SAVE_RESULT) {
// int w = cap.get(cv::CAP_PROP_FRAME_WIDTH);
// int h = cap.get(cv::CAP_PROP_FRAME_HEIGHT);
// resultVideo.open("result.mp4", cv::VideoWriter::fourcc('M','P','4','V'), 30, cv::Size(w, h));
// }
cv::Mat frame; cv::Mat frame;
cv::Mat frame_crop; cv::Mat frame_crop;
cv::Mat dnn_input; cv::Mat dnn_input;
@@ -546,6 +555,9 @@ int main(int argc, char *argv[])
{ {
yolo[i].init(par.net); yolo[i].init(par.net);
yolo[i].thresh = 0.25; yolo[i].thresh = 0.25;
// if(SAVE_RESULT)
// resultVideo << frame;
} }
// tk::dnn::Yolo3Detection yolo; // tk::dnn::Yolo3Detection yolo;
// yolo.init(net); // yolo.init(net);
include/Layer.h → include/tkDNN/Layer.h
+125 -40
View File
@@ -1,7 +1,8 @@
#ifndef LAYER_H #ifndef LAYER_H
#define LAYER_H #define LAYER_H
#include <iostream> #include<iostream>
#include<vector>
#include "utils.h" #include "utils.h"
#include "Network.h" #include "Network.h"
@@ -10,10 +11,11 @@ namespace tk
namespace dnn namespace dnn
{ {
enum layerType_t enum layerType_t {
{ LAYER_INPUT,
LAYER_DENSE, LAYER_DENSE,
LAYER_CONV2D, LAYER_CONV2D,
LAYER_LSTM,
LAYER_ACTIVATION, LAYER_ACTIVATION,
LAYER_FLATTEN, LAYER_FLATTEN,
LAYER_MULADD, LAYER_MULADD,
@@ -52,36 +54,23 @@ public:
std::string getLayerName() std::string getLayerName()
{ {
layerType_t type = getLayerType(); layerType_t type = getLayerType();
switch (type) switch(type) {
{ case LAYER_INPUT: return "Input";
case LAYER_DENSE: case LAYER_DENSE: return "Dense";
return "Dense"; case LAYER_CONV2D: return "Conv2d";
case LAYER_CONV2D: case LAYER_LSTM: return "LSTM";
return "Conv2d"; case LAYER_ACTIVATION: return "Activation";
case LAYER_ACTIVATION: case LAYER_FLATTEN: return "Flatten";
return "Activation"; case LAYER_MULADD: return "MulAdd";
case LAYER_FLATTEN: case LAYER_POOLING: return "Pooling";
return "Flatten"; case LAYER_SOFTMAX: return "Softmax";
case LAYER_MULADD: case LAYER_ROUTE: return "Route";
return "MulAdd"; case LAYER_REORG: return "Reorg";
case LAYER_POOLING: case LAYER_SHORTCUT: return "Shortcut";
return "Pooling"; case LAYER_UPSAMPLE: return "Upsample";
case LAYER_SOFTMAX: case LAYER_REGION: return "Region";
return "Softmax"; case LAYER_YOLO: return "Yolo";
case LAYER_ROUTE: default: return "unknown";
return "Route";
case LAYER_REORG:
return "Reorg";
case LAYER_SHORTCUT:
return "Shortcut";
case LAYER_UPSAMPLE:
return "Upsample";
case LAYER_REGION:
return "Region";
case LAYER_YOLO:
return "Yolo";
default:
return "unknown";
} }
} }
@@ -98,7 +87,7 @@ class LayerWgs : public Layer
public: public:
LayerWgs(Network *net, int inputs, int outputs, int kh, int kw, int kt, LayerWgs(Network *net, int inputs, int outputs, int kh, int kw, int kt,
const char *fname_weights, bool batchnorm = false); std::string fname_weights, bool batchnorm = false);
virtual ~LayerWgs(); virtual ~LayerWgs();
int inputs, outputs; int inputs, outputs;
@@ -124,6 +113,27 @@ public:
__half *variance16_h, *variance16_d; __half *variance16_h, *variance16_d;
}; };
/**
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) {
return dstData;
}
};
/** /**
Dense (full interconnection) layer Dense (full interconnection) layer
*/ */
@@ -131,7 +141,7 @@ class Dense : public LayerWgs
{ {
public: public:
Dense(Network *net, int out_ch, const char *fname_weights); Dense(Network *net, int out_ch, std::string fname_weights);
virtual ~Dense(); virtual ~Dense();
virtual layerType_t getLayerType() { return LAYER_DENSE; }; virtual layerType_t getLayerType() { return LAYER_DENSE; };
@@ -168,14 +178,22 @@ protected:
/** /**
Convolutional 2D layer Convolutional 2D layer
WEIGHTS shape: OUTCH, INCH, KH, KW ...
BIAS shape: OUTCH
with BATCHNORM:
scales: OUTCH
means: OUTCH
variance: OUTCH
*/ */
class Conv2d : public LayerWgs class Conv2d : public LayerWgs
{ {
public: public:
Conv2d(Network *net, int out_ch, int kernelH, int kernelW, Conv2d( Network *net, int out_ch, int kernelH, int kernelW,
int strideH, int strideW, int paddingH, int paddingW, int strideH, int strideW, int paddingH, int paddingW,
const char *fname_weights, bool batchnorm = false); std::string fname_weights, bool batchnorm = false);
virtual ~Conv2d(); virtual ~Conv2d();
virtual layerType_t getLayerType() { return LAYER_CONV2D; }; virtual layerType_t getLayerType() { return LAYER_CONV2D; };
@@ -193,6 +211,72 @@ protected:
size_t ws_sizeInBytes; 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<cudnnTensorDescriptor_t> 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
};
/** /**
Flatten layer Flatten layer
is actually a matrix transposition is actually a matrix transposition
@@ -384,13 +468,14 @@ public:
int sort_class; int sort_class;
}; };
Yolo(Network *net, int classes, int num, const char *fname_weights); Yolo(Network *net, int classes, int num, std::string fname_weights);
virtual ~Yolo(); virtual ~Yolo();
virtual layerType_t getLayerType() { return LAYER_YOLO; }; virtual layerType_t getLayerType() { return LAYER_YOLO; };
int classes, num; int classes, num;
dnnType *mask_h, *mask_d; //anchors dnnType *mask_h, *mask_d; //anchors
dnnType *bias_h, *bias_d; //anchors dnnType *bias_h, *bias_d; //anchors
std::vector<std::string> classesNames;
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData); virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
int computeDetections(Yolo::detection *dets, int &ndets, int netw, int neth, float thresh); int computeDetections(Yolo::detection *dets, int &ndets, int netw, int neth, float thresh);
@@ -423,7 +508,7 @@ class RegionInterpret
public: public:
RegionInterpret(dataDim_t input_dim, dataDim_t output_dim, RegionInterpret(dataDim_t input_dim, dataDim_t output_dim,
int classes, int coords, int num, float thresh, const char *fname_weights); int classes, int coords, int num, float thresh, std::string fname_weights);
~RegionInterpret(); ~RegionInterpret();
dataDim_t input_dim, output_dim; dataDim_t input_dim, output_dim;
include/Network.h → include/tkDNN/Network.h
View File
include/NetworkRT.h → include/tkDNN/NetworkRT.h
View File
include/Yolo3Detection.h → include/tkDNN/Yolo3Detection.h
+12
View File
@@ -46,6 +46,9 @@ public:
// this is filled with results // this is filled with results
std::vector<tk::dnn::box> detected; std::vector<tk::dnn::box> detected;
// keep track of inference times (ms)
std::vector<double> stats;
Yolo3Detection() {} Yolo3Detection() {}
virtual ~Yolo3Detection() {} virtual ~Yolo3Detection() {}
@@ -58,6 +61,15 @@ public:
bool init(std::string tensor_path); bool init(std::string tensor_path);
void addBorders(cv::Mat &imageORIG, cv::Mat &imageWBorders, int &top, int &left); void addBorders(cv::Mat &imageORIG, cv::Mat &imageWBorders, int &top, int &left);
void update(cv::Mat &frame); void update(cv::Mat &frame);
tk::dnn::Yolo* getYoloLayer(int n=0) {
if(n<3)
return yolo[n];
else
return nullptr;
}
}; };
} // namespace dnn } // namespace dnn
include/kernels.h → include/tkDNN/kernels.h
View File
include/tkDNN/models/Yolo3.h
+289
View File
@@ -0,0 +1,289 @@
int preYoloFilters = (classes+5)*3;
std::string input_bin = bin_path + "/layers/input.bin";
std::vector<std::string> output_bins = {
bin_path + "/debug/layer82_out.bin",
bin_path + "/debug/layer94_out.bin",
bin_path + "/debug/layer106_out.bin"
};
std::string c0_bin = bin_path + "/layers/c0.bin";
std::string c1_bin = bin_path + "/layers/c1.bin";
std::string c2_bin = bin_path + "/layers/c2.bin";
std::string c3_bin = bin_path + "/layers/c3.bin";
std::string c5_bin = bin_path + "/layers/c5.bin";
std::string c6_bin = bin_path + "/layers/c6.bin";
std::string c7_bin = bin_path + "/layers/c7.bin";
std::string c9_bin = bin_path + "/layers/c9.bin";
std::string c10_bin = bin_path + "/layers/c10.bin";
std::string c12_bin = bin_path + "/layers/c12.bin";
std::string c13_bin = bin_path + "/layers/c13.bin";
std::string c14_bin = bin_path + "/layers/c14.bin";
std::string c16_bin = bin_path + "/layers/c16.bin";
std::string c17_bin = bin_path + "/layers/c17.bin";
std::string c19_bin = bin_path + "/layers/c19.bin";
std::string c20_bin = bin_path + "/layers/c20.bin";
std::string c22_bin = bin_path + "/layers/c22.bin";
std::string c23_bin = bin_path + "/layers/c23.bin";
std::string c25_bin = bin_path + "/layers/c25.bin";
std::string c26_bin = bin_path + "/layers/c26.bin";
std::string c28_bin = bin_path + "/layers/c28.bin";
std::string c29_bin = bin_path + "/layers/c29.bin";
std::string c31_bin = bin_path + "/layers/c31.bin";
std::string c32_bin = bin_path + "/layers/c32.bin";
std::string c34_bin = bin_path + "/layers/c34.bin";
std::string c35_bin = bin_path + "/layers/c35.bin";
std::string c37_bin = bin_path + "/layers/c37.bin";
std::string c38_bin = bin_path + "/layers/c38.bin";
std::string c39_bin = bin_path + "/layers/c39.bin";
std::string c41_bin = bin_path + "/layers/c41.bin";
std::string c42_bin = bin_path + "/layers/c42.bin";
std::string c44_bin = bin_path + "/layers/c44.bin";
std::string c45_bin = bin_path + "/layers/c45.bin";
std::string c47_bin = bin_path + "/layers/c47.bin";
std::string c48_bin = bin_path + "/layers/c48.bin";
std::string c50_bin = bin_path + "/layers/c50.bin";
std::string c51_bin = bin_path + "/layers/c51.bin";
std::string c53_bin = bin_path + "/layers/c53.bin";
std::string c54_bin = bin_path + "/layers/c54.bin";
std::string c56_bin = bin_path + "/layers/c56.bin";
std::string c57_bin = bin_path + "/layers/c57.bin";
std::string c59_bin = bin_path + "/layers/c59.bin";
std::string c60_bin = bin_path + "/layers/c60.bin";
std::string c62_bin = bin_path + "/layers/c62.bin";
std::string c63_bin = bin_path + "/layers/c63.bin";
std::string c64_bin = bin_path + "/layers/c64.bin";
std::string c66_bin = bin_path + "/layers/c66.bin";
std::string c67_bin = bin_path + "/layers/c67.bin";
std::string c69_bin = bin_path + "/layers/c69.bin";
std::string c70_bin = bin_path + "/layers/c70.bin";
std::string c72_bin = bin_path + "/layers/c72.bin";
std::string c73_bin = bin_path + "/layers/c73.bin";
std::string c75_bin = bin_path + "/layers/c75.bin";
std::string c76_bin = bin_path + "/layers/c76.bin";
std::string c77_bin = bin_path + "/layers/c77.bin";
std::string c78_bin = bin_path + "/layers/c78.bin";
std::string c79_bin = bin_path + "/layers/c79.bin";
std::string c80_bin = bin_path + "/layers/c80.bin";
std::string c81_bin = bin_path + "/layers/c81.bin";
std::string g82_bin = bin_path + "/layers/g82.bin";
std::string c84_bin = bin_path + "/layers/c84.bin";
std::string c87_bin = bin_path + "/layers/c87.bin";
std::string c88_bin = bin_path + "/layers/c88.bin";
std::string c89_bin = bin_path + "/layers/c89.bin";
std::string c90_bin = bin_path + "/layers/c90.bin";
std::string c91_bin = bin_path + "/layers/c91.bin";
std::string c92_bin = bin_path + "/layers/c92.bin";
std::string c93_bin = bin_path + "/layers/c93.bin";
std::string g94_bin = bin_path + "/layers/g94.bin";
std::string c96_bin = bin_path + "/layers/c96.bin";
std::string c99_bin = bin_path + "/layers/c99.bin";
std::string c100_bin = bin_path + "/layers/c100.bin";
std::string c101_bin = bin_path + "/layers/c101.bin";
std::string c102_bin = bin_path + "/layers/c102.bin";
std::string c103_bin = bin_path + "/layers/c103.bin";
std::string c104_bin = bin_path + "/layers/c104.bin";
std::string c105_bin = bin_path + "/layers/c105.bin";
std::string g106_bin = bin_path + "/layers/g106.bin";
tk::dnn::Conv2d c0 (&net, 32, 3, 3, 1, 1, 1, 1, c0_bin, true);
tk::dnn::Activation a0 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c1 (&net, 64, 3, 3, 2, 2, 1, 1, c1_bin, true);
tk::dnn::Activation a1 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c2 (&net, 32, 1, 1, 1, 1, 0, 0, c2_bin, true);
tk::dnn::Activation a2 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c3 (&net, 64, 3, 3, 1, 1, 1, 1, c3_bin, true);
tk::dnn::Activation a3 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s4 (&net, &a1);
tk::dnn::Conv2d c5 (&net, 128, 3, 3, 2, 2, 1, 1, c5_bin, true);
tk::dnn::Activation a5 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c6 (&net, 64, 1, 1, 1, 1, 0, 0, c6_bin, true);
tk::dnn::Activation a6 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c7 (&net, 128, 3, 3, 1, 1, 1, 1, c7_bin, true);
tk::dnn::Activation a7 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s8 (&net, &a5);
tk::dnn::Conv2d c9 (&net, 64, 1, 1, 1, 1, 0, 0, c9_bin, true);
tk::dnn::Activation a9 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c10 (&net, 128, 3, 3, 1, 1, 1, 1, c10_bin, true);
tk::dnn::Activation a10 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s11 (&net, &s8);
tk::dnn::Conv2d c12 (&net, 256, 3, 3, 2, 2, 1, 1, c12_bin, true);
tk::dnn::Activation a12 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c13 (&net, 128, 1, 1, 1, 1, 0, 0, c13_bin, true);
tk::dnn::Activation a13 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c14 (&net, 256, 3, 3, 1, 1, 1, 1, c14_bin, true);
tk::dnn::Activation a14 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s15 (&net, &a12);
tk::dnn::Conv2d c16 (&net, 128, 1, 1, 1, 1, 0, 0, c16_bin, true);
tk::dnn::Activation a16 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c17 (&net, 256, 3, 3, 1, 1, 1, 1, c17_bin, true);
tk::dnn::Activation a17 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s18 (&net, &s15);
tk::dnn::Conv2d c19 (&net, 128, 1, 1, 1, 1, 0, 0, c19_bin, true);
tk::dnn::Activation a19 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c20 (&net, 256, 3, 3, 1, 1, 1, 1, c20_bin, true);
tk::dnn::Activation a20 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s21 (&net, &s18);
tk::dnn::Conv2d c22 (&net, 128, 1, 1, 1, 1, 0, 0, c22_bin, true);
tk::dnn::Activation a22 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c23 (&net, 256, 3, 3, 1, 1, 1, 1, c23_bin, true);
tk::dnn::Activation a23 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s24 (&net, &s21);
tk::dnn::Conv2d c25 (&net, 128, 1, 1, 1, 1, 0, 0, c25_bin, true);
tk::dnn::Activation a25 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c26 (&net, 256, 3, 3, 1, 1, 1, 1, c26_bin, true);
tk::dnn::Activation a26 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s27 (&net, &s24);
tk::dnn::Conv2d c28 (&net, 128, 1, 1, 1, 1, 0, 0, c28_bin, true);
tk::dnn::Activation a28 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c29 (&net, 256, 3, 3, 1, 1, 1, 1, c29_bin, true);
tk::dnn::Activation a29 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s30 (&net, &s27);
tk::dnn::Conv2d c31 (&net, 128, 1, 1, 1, 1, 0, 0, c31_bin, true);
tk::dnn::Activation a31 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c32 (&net, 256, 3, 3, 1, 1, 1, 1, c32_bin, true);
tk::dnn::Activation a32 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s33 (&net, &s30);
tk::dnn::Conv2d c34 (&net, 128, 1, 1, 1, 1, 0, 0, c34_bin, true);
tk::dnn::Activation a34 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c35 (&net, 256, 3, 3, 1, 1, 1, 1, c35_bin, true);
tk::dnn::Activation a35 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s36 (&net, &s33);
tk::dnn::Conv2d c37 (&net, 512, 3, 3, 2, 2, 1, 1, c37_bin, true);
tk::dnn::Activation a37 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c38 (&net, 256, 1, 1, 1, 1, 0, 0, c38_bin, true);
tk::dnn::Activation a38 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c39 (&net, 512, 3, 3, 1, 1, 1, 1, c39_bin, true);
tk::dnn::Activation a39 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s40 (&net, &a37);
tk::dnn::Conv2d c41 (&net, 256, 1, 1, 1, 1, 0, 0, c41_bin, true);
tk::dnn::Activation a41 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c42 (&net, 512, 3, 3, 1, 1, 1, 1, c42_bin, true);
tk::dnn::Activation a42 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s43 (&net, &s40);
tk::dnn::Conv2d c44 (&net, 256, 1, 1, 1, 1, 0, 0, c44_bin, true);
tk::dnn::Activation a44 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c45 (&net, 512, 3, 3, 1, 1, 1, 1, c45_bin, true);
tk::dnn::Activation a45 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s46 (&net, &s43);
tk::dnn::Conv2d c47 (&net, 256, 1, 1, 1, 1, 0, 0, c47_bin, true);
tk::dnn::Activation a47 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c48 (&net, 512, 3, 3, 1, 1, 1, 1, c48_bin, true);
tk::dnn::Activation a48 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s49 (&net, &s46);
tk::dnn::Conv2d c50 (&net, 256, 1, 1, 1, 1, 0, 0, c50_bin, true);
tk::dnn::Activation a50 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c51 (&net, 512, 3, 3, 1, 1, 1, 1, c51_bin, true);
tk::dnn::Activation a51 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s52 (&net, &s49);
tk::dnn::Conv2d c53 (&net, 256, 1, 1, 1, 1, 0, 0, c53_bin, true);
tk::dnn::Activation a53 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c54 (&net, 512, 3, 3, 1, 1, 1, 1, c54_bin, true);
tk::dnn::Activation a54 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s55 (&net, &s52);
tk::dnn::Conv2d c56 (&net, 256, 1, 1, 1, 1, 0, 0, c56_bin, true);
tk::dnn::Activation a56 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c57 (&net, 512, 3, 3, 1, 1, 1, 1, c57_bin, true);
tk::dnn::Activation a57 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s58 (&net, &s55);
tk::dnn::Conv2d c59 (&net, 256, 1, 1, 1, 1, 0, 0, c59_bin, true);
tk::dnn::Activation a59 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c60 (&net, 512, 3, 3, 1, 1, 1, 1, c60_bin, true);
tk::dnn::Activation a60 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s61 (&net, &s58);
tk::dnn::Conv2d c62 (&net,1024, 3, 3, 2, 2, 1, 1, c62_bin, true);
tk::dnn::Activation a62 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c63 (&net, 512, 1, 1, 1, 1, 0, 0, c63_bin, true);
tk::dnn::Activation a63 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c64 (&net,1024, 3, 3, 1, 1, 1, 1, c64_bin, true);
tk::dnn::Activation a64 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s65 (&net, &a62);
tk::dnn::Conv2d c66 (&net, 512, 1, 1, 1, 1, 0, 0, c66_bin, true);
tk::dnn::Activation a66 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c67 (&net,1024, 3, 3, 1, 1, 1, 1, c67_bin, true);
tk::dnn::Activation a67 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s68 (&net, &s65);
tk::dnn::Conv2d c69 (&net, 512, 1, 1, 1, 1, 0, 0, c69_bin, true);
tk::dnn::Activation a69 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c70 (&net,1024, 3, 3, 1, 1, 1, 1, c70_bin, true);
tk::dnn::Activation a70 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s71 (&net, &s68);
tk::dnn::Conv2d c72 (&net, 512, 1, 1, 1, 1, 0, 0, c72_bin, true);
tk::dnn::Activation a72 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c73 (&net,1024, 3, 3, 1, 1, 1, 1, c73_bin, true);
tk::dnn::Activation a73 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s74 (&net, &s71);
tk::dnn::Conv2d c75 (&net, 512, 1, 1, 1, 1, 0, 0, c75_bin, true);
tk::dnn::Activation a75 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c76 (&net,1024, 3, 3, 1, 1, 1, 1, c76_bin, true);
tk::dnn::Activation a76 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c77 (&net, 512, 1, 1, 1, 1, 0, 0, c77_bin, true);
tk::dnn::Activation a77 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c78 (&net,1024, 3, 3, 1, 1, 1, 1, c78_bin, true);
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, preYoloFilters, 1, 1, 1, 1, 0, 0, c81_bin, false);
tk::dnn::Yolo yolo0 (&net, classes, 3, g82_bin);
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::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);
tk::dnn::Activation a88 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c89 (&net, 256, 1, 1, 1, 1, 0, 0, c89_bin, true);
tk::dnn::Activation a89 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c90 (&net, 512, 3, 3, 1, 1, 1, 1, c90_bin, true);
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, preYoloFilters, 1, 1, 1, 1, 0, 0, c93_bin, false);
tk::dnn::Yolo yolo1 (&net, classes, 3, g94_bin);
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::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);
tk::dnn::Activation a100 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c101 (&net, 128, 1, 1, 1, 1, 0, 0, c101_bin, true);
tk::dnn::Activation a101 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c102 (&net, 256, 3, 3, 1, 1, 1, 1, c102_bin, true);
tk::dnn::Activation a102 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c103 (&net, 128, 1, 1, 1, 1, 0, 0, c103_bin, true);
tk::dnn::Activation a103 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c104 (&net, 256, 3, 3, 1, 1, 1, 1, c104_bin, true);
tk::dnn::Activation a104 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c105 (&net, preYoloFilters, 1, 1, 1, 1, 0, 0, c105_bin, false);
tk::dnn::Yolo yolo2 (&net, classes, 3, g106_bin);
yolo[0] = &yolo0;
yolo[1] = &yolo1;
yolo[2] = &yolo2;
include/pluginsRT/ActivationLeakyRT.h → include/tkDNN/pluginsRT/ActivationLeakyRT.h
View File
include/pluginsRT/Int8Calibrator.h → include/tkDNN/pluginsRT/Int8Calibrator.h
View File
include/pluginsRT/RegionRT.h → include/tkDNN/pluginsRT/RegionRT.h
View File
include/pluginsRT/ReorgRT.h → include/tkDNN/pluginsRT/ReorgRT.h
View File
include/pluginsRT/RouteRT.h → include/tkDNN/pluginsRT/RouteRT.h
View File
include/pluginsRT/ShortcutRT.h → include/tkDNN/pluginsRT/ShortcutRT.h
View File
include/pluginsRT/UpsampleRT.h → include/tkDNN/pluginsRT/UpsampleRT.h
View File
include/pluginsRT/YoloRT.h → include/tkDNN/pluginsRT/YoloRT.h
+14 -1
View File
@@ -1,6 +1,8 @@
#include<cassert> #include<cassert>
#include "../kernels.h" #include "../kernels.h"
#define YOLORT_CLASSNAME_W 256
class YoloRT : public IPlugin { class YoloRT : public IPlugin {
@@ -16,6 +18,7 @@ public:
if(yolo != nullptr) { if(yolo != nullptr) {
memcpy(mask, yolo->mask_h, sizeof(dnnType)*num); memcpy(mask, yolo->mask_h, sizeof(dnnType)*num);
memcpy(bias, yolo->bias_h, sizeof(dnnType)*num*3*2); memcpy(bias, yolo->bias_h, sizeof(dnnType)*num*3*2);
classesNames = yolo->classesNames;
} }
} }
@@ -72,7 +75,7 @@ public:
virtual size_t getSerializationSize() override { virtual size_t getSerializationSize() override {
return 5*sizeof(int) + num*sizeof(dnnType) + num*3*2*sizeof(dnnType); return 5*sizeof(int) + num*sizeof(dnnType) + num*3*2*sizeof(dnnType) + YOLORT_CLASSNAME_W*classes*sizeof(char);
} }
virtual void serialize(void* buffer) override { virtual void serialize(void* buffer) override {
@@ -86,10 +89,20 @@ public:
tk::dnn::writeBUF(buf, mask[i]); tk::dnn::writeBUF(buf, mask[i]);
for(int i=0; i<3*2*num; i++) for(int i=0; i<3*2*num; i++)
tk::dnn::writeBUF(buf, bias[i]); tk::dnn::writeBUF(buf, bias[i]);
// save classes names
for(int i=0; i<classes; i++) {
char tmp[YOLORT_CLASSNAME_W];
strcpy(tmp, classesNames[i].c_str());
for(int j=0; j<YOLORT_CLASSNAME_W; j++) {
tk::dnn::writeBUF(buf, tmp[j]);
}
}
} }
int c, h, w; int c, h, w;
int classes, num; int classes, num;
std::vector<std::string> classesNames;
dnnType *mask; dnnType *mask;
dnnType *bias; dnnType *bias;
include/tkdnn.h → include/tkDNN/tkdnn.h
+1 -1
View File
@@ -5,4 +5,4 @@
#include "Layer.h" #include "Layer.h"
#include "NetworkRT.h" #include "NetworkRT.h"
#define TKDNN_VERSION 300 #define TKDNN_VERSION 400
include/utils.h → include/tkDNN/utils.h
+3 -3
View File
@@ -104,9 +104,9 @@
void printCenteredTitle(const char *title, char fill, int dim); void printCenteredTitle(const char *title, char fill, int dim);
bool fileExist(const char *fname); bool fileExist(const char *fname);
void readBinaryFile(const char *fname, int size, dnnType **data_h, dnnType **data_d, int seek = 0); 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 printDeviceVector(int size, dnnType* vec_d, bool device = true);
void resize(int size, dnnType **data); void resize(int size, dnnType **data);
void matrixTranspose(cublasHandle_t handle, dnnType *srcData, dnnType *dstData, int rows, int cols); void matrixTranspose(cublasHandle_t handle, dnnType *srcData, dnnType *dstData, int rows, int cols);
src/Conv2d.cpp
+5 -6
View File
@@ -7,13 +7,12 @@ namespace tk
namespace dnn namespace dnn
{ {
Conv2d::Conv2d(Network *net, int out_ch, int kernelH, int kernelW, Conv2d::Conv2d( Network *net, int out_ch, int kernelH, int kernelW,
int strideH, int strideW, int paddingH, int paddingW, int strideH, int strideW, int paddingH, int paddingW,
const char *fname_weights, bool batchnorm) : std::string fname_weights, bool batchnorm) :
LayerWgs(net, net->getOutputDim().c, out_ch, kernelH, kernelW, 1, LayerWgs(net, net->getOutputDim().c, out_ch, kernelH, kernelW, 1,
fname_weights, batchnorm) fname_weights, batchnorm) {
{
this->kernelH = kernelH; this->kernelH = kernelH;
this->kernelW = kernelW; this->kernelW = kernelW;
src/Dense.cpp
+2 -2
View File
@@ -7,8 +7,8 @@ namespace tk
namespace dnn namespace dnn
{ {
Dense::Dense(Network *net, int out_ch, const char *fname_weights) : LayerWgs(net, net->getOutputDim().tot(), out_ch, 1, 1, 1, fname_weights) Dense::Dense(Network *net, int out_ch, std::string fname_weights) :
{ LayerWgs(net, net->getOutputDim().tot(), out_ch, 1, 1, 1, fname_weights) {
output_dim.n = 1; output_dim.n = 1;
output_dim.c = out_ch; output_dim.c = out_ch;
src/LSTM.cpp
+327
View File
@@ -0,0 +1,327 @@
#include <iostream>
#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<cudnnTensorDescriptor_t> x_vec(seqLen);
std::vector<cudnnTensorDescriptor_t> 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: "<<cudnn_params << ", bytes: "<<cudnn_param_size<<"\n";
// Set param descriptors
checkCUDNN(cudnnCreateFilterDescriptor(&w_desc_));
int dim_w[3] = {1, 1, 1};
dim_w[0] = cudnn_params;
checkCUDNN(cudnnSetFilterNdDescriptor(w_desc_,
net->dataType, net->tensorFormat, 3, dim_w));
// load params
std::cout<<"Reading weights: PARAMS="<<cudnn_params*2<<"\n";
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: "<<wf_ptr<<" wb "<<wb_ptr<<"\n";
// set output dim
output_dim = input_dim;
output_dim.c = stateSize*(bidirectional ? 2 : 1);
// if retunseq is disabled only the last timestep is returned
if(!returnSeq) {
output_dim.h = 1;
output_dim.w = 1;
}
//allocate data for infer result
checkCuda( cudaMalloc(&dstData, output_dim.tot()*sizeof(dnnType)) );
// used during inference
one_output_dim = input_dim;
one_output_dim.c = stateSize;
checkCuda( cudaMalloc(&srcF, input_dim.tot()*sizeof(dnnType)) );
checkCuda( cudaMalloc(&srcB, input_dim.tot()*sizeof(dnnType)) );
checkCuda( cudaMalloc(&dstF, one_output_dim.tot()*sizeof(dnnType)) );
checkCuda( cudaMalloc(&dstB_NR, one_output_dim.tot()*sizeof(dnnType)) );
checkCuda( cudaMalloc(&dstB, one_output_dim.tot()*sizeof(dnnType)) );
/*
// Query weight layout
cudnnFilterDescriptor_t m_desc;
checkCUDNN(cudnnCreateFilterDescriptor(&m_desc));
dnnType *p;
int n = 8; // lstm layers
printCenteredTitle("WEIGHTS", '=', 20);
for (int i = 0; i < numLayers; ++i) {
for (int j = 0; j < n; ++j) {
checkCUDNN(cudnnGetRNNLinLayerMatrixParams(net->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-> "<<tot<<"\n\n";
}
}
printCenteredTitle("BIAS", '=', 20);
for (int i = 0; i < numLayers; ++i) {
for (int j = 0; j < n; ++j) {
checkCUDNN(cudnnGetRNNLinLayerBiasParams(net->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-> "<<tot<<"\n\n";
}
}
checkCUDNN(cudnnDestroyFilterDescriptor(m_desc));
*/
}
LSTM::~LSTM() {
checkCuda(cudaFree(hx_ptr));
checkCuda(cudaFree(cx_ptr));
checkCuda(cudaFree(hy_ptr));
checkCuda(cudaFree(cy_ptr));
checkCuda(cudaFree(w_ptr ));
checkCuda(cudaFree(work_space_ ));
checkCuda(cudaFree(dropout_states_));
checkCuda(cudaFree(srcF));
checkCuda(cudaFree(srcB));
checkCuda(cudaFree(dstF));
checkCuda(cudaFree(dstB_NR));
checkCuda(cudaFree(dstB));
checkCuda(cudaFree(dstData));
}
dnnType* LSTM::infer(dataDim_t &dim, dnnType* srcData) {
// transpose input
matrixTranspose(net->cublasHandle, srcData, srcF, dim.c, dim.h*dim.w*dim.l);
// build srcB as reversed srcF
for(int i=0; i<input_dim.w; i++) {
int off_0 = i*(input_dim.c);
int off_1 = (i+1)*(input_dim.c);
checkCuda( cudaMemcpy(srcB + dim.tot() - off_1, srcF + off_0,
input_dim.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
}
// forward
{
// 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)
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; i<one_output_dim.w; i++) {
int off_0 = i*(one_output_dim.c);
int off_1 = (i+1)*(one_output_dim.c);
checkCuda( cudaMemcpy(dstB + one_output_dim.tot() - off_1, dstB_NR + off_0,
one_output_dim.c*sizeof(dnnType), cudaMemcpyDeviceToDevice));
}
// if retunseq is disabled only the last timestep is returned
if(returnSeq) {
// forward transpose
matrixTranspose(net->cublasHandle, 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;
}
}}
src/LayerWgs.cpp
+1 -2
View File
@@ -11,8 +11,7 @@ namespace dnn
LayerWgs::LayerWgs(Network *net, int inputs, int outputs, LayerWgs::LayerWgs(Network *net, int inputs, int outputs,
int kh, int kw, int kl, int kh, int kw, int kl,
const char *fname_weights, bool batchnorm) : Layer(net) std::string fname_weights, bool batchnorm) : Layer(net) {
{
this->inputs = inputs; this->inputs = inputs;
this->outputs = outputs; this->outputs = outputs;
src/NetworkRT.cpp
+28 -5
View File
@@ -35,7 +35,7 @@ NetworkRT::NetworkRT(Network *net, const char *name) {
builderRT = createInferBuilder(loggerRT); builderRT = createInferBuilder(loggerRT);
std::cout<<"Float16 support: "<<builderRT->platformHasFastFp16()<<"\n"; std::cout<<"Float16 support: "<<builderRT->platformHasFastFp16()<<"\n";
std::cout<<"Int8 support: "<<builderRT->platformHasFastInt8()<<"\n"; std::cout<<"Int8 support: "<<builderRT->platformHasFastInt8()<<"\n";
//std::cout<<"DLAs: "<<builderRT->getNbDLACores()<<"\n"; std::cout<<"DLAs: "<<builderRT->getNbDLACores()<<"\n";
networkRT = builderRT->createNetwork(); networkRT = builderRT->createNetwork();
if(!fileExist(name)) { if(!fileExist(name)) {
@@ -51,7 +51,6 @@ NetworkRT::NetworkRT(Network *net, const char *name) {
dtRT = DataType::kHALF; dtRT = DataType::kHALF;
builderRT->setHalf2Mode(true); builderRT->setHalf2Mode(true);
} }
/*
if(net->dla && builderRT->getNbDLACores() > 0) { if(net->dla && builderRT->getNbDLACores() > 0) {
dtRT = DataType::kHALF; dtRT = DataType::kHALF;
builderRT->setFp16Mode(true); builderRT->setFp16Mode(true);
@@ -59,7 +58,6 @@ NetworkRT::NetworkRT(Network *net, const char *name) {
builderRT->setDefaultDeviceType(DeviceType::kDLA); builderRT->setDefaultDeviceType(DeviceType::kDLA);
builderRT->setDLACore(0); builderRT->setDLACore(0);
} }
*/
//add input layer //add input layer
ITensor *input = networkRT->addInput("data", DataType::kFLOAT, ITensor *input = networkRT->addInput("data", DataType::kFLOAT,
@@ -276,10 +274,19 @@ ILayer* NetworkRT::convert_layer(ITensor *input, Activation *l) {
if(l->act_mode == ACTIVATION_LEAKY) { if(l->act_mode == ACTIVATION_LEAKY) {
//std::cout<<"New plugin LEAKY\n"; //std::cout<<"New plugin LEAKY\n";
#if NV_TENSORRT_MAJOR < 6
// plugin version
IPlugin *plugin = new ActivationLeakyRT(); IPlugin *plugin = new ActivationLeakyRT();
IPluginLayer *lRT = networkRT->addPlugin(&input, 1, *plugin); IPluginLayer *lRT = networkRT->addPlugin(&input, 1, *plugin);
checkNULL(lRT); checkNULL(lRT);
return lRT; return lRT;
#else
IActivationLayer *lRT = networkRT->addActivation(*input, ActivationType::kLEAKY_RELU);
lRT->setAlpha(0.1);
checkNULL(lRT);
return lRT;
#endif
} else if(l->act_mode == CUDNN_ACTIVATION_RELU) { } else if(l->act_mode == CUDNN_ACTIVATION_RELU) {
IActivationLayer *lRT = networkRT->addActivation(*input, ActivationType::kRELU); IActivationLayer *lRT = networkRT->addActivation(*input, ActivationType::kRELU);
@@ -340,14 +347,21 @@ ILayer* NetworkRT::convert_layer(ITensor *input, Shortcut *l) {
//std::cout<<"convert Shortcut\n"; //std::cout<<"convert Shortcut\n";
//std::cout<<"New plugin Shortcut\n"; //std::cout<<"New plugin Shortcut\n";
ITensor *back_tens = tensors[l->backLayer];
IPlugin *plugin = new ShortcutRT();
ITensor *back_tens = tensors[l->backLayer];
/*
// plugin version
IPlugin *plugin = new ShortcutRT();
ITensor **inputs = new ITensor*[2]; ITensor **inputs = new ITensor*[2];
inputs[0] = input; inputs[0] = input;
inputs[1] = back_tens; inputs[1] = back_tens;
IPluginLayer *lRT = networkRT->addPlugin(inputs, 2, *plugin); IPluginLayer *lRT = networkRT->addPlugin(inputs, 2, *plugin);
checkNULL(lRT); checkNULL(lRT);
*/
IElementWiseLayer *lRT = networkRT->addElementWise(*input, *back_tens, ElementWiseOperation::kSUM);
checkNULL(lRT);
return lRT; return lRT;
} }
@@ -461,6 +475,15 @@ IPlugin* PluginFactory::createPlugin(const char* layerName, const void* serialDa
for(int i=0; i<3*2*r->num; i++) for(int i=0; i<3*2*r->num; i++)
r->bias[i] = readBUF<dnnType>(buf); r->bias[i] = readBUF<dnnType>(buf);
// save classes names
r->classesNames.resize(r->classes);
for(int i=0; i<r->classes; i++) {
char tmp[YOLORT_CLASSNAME_W];
for(int j=0; j<YOLORT_CLASSNAME_W; j++)
tmp[j] = readBUF<char>(buf);
r->classesNames[i] = std::string(tmp);
}
yolos[n_yolos++] = r; yolos[n_yolos++] = r;
return r; return r;
} }
src/Region.cpp
+1 -1
View File
@@ -66,7 +66,7 @@ dnnType* Region::infer(dataDim_t &dim, dnnType* srcData) {
/* Intepret class */ /* Intepret class */
RegionInterpret::RegionInterpret(dataDim_t input_dim, dataDim_t output_dim, RegionInterpret::RegionInterpret(dataDim_t input_dim, dataDim_t output_dim,
int classes, int coords, int num, float thresh, const char* fname_weights) { int classes, int coords, int num, float thresh, std::string fname_weights) {
this->input_dim = input_dim; this->input_dim = input_dim;
this->output_dim = output_dim; this->output_dim = output_dim;
src/Yolo.cpp
+8 -2
View File
@@ -11,20 +11,26 @@
namespace tk { namespace dnn { namespace tk { namespace dnn {
Yolo::Yolo(Network *net, int classes, int num, const char* fname_weights) : Yolo::Yolo(Network *net, int classes, int num, std::string fname_weights) :
Layer(net) { Layer(net) {
this->classes = classes; this->classes = classes;
this->num = num; this->num = num;
// load anchors // load anchors
if(fname_weights != nullptr) { if(fname_weights != "") {
int seek = 0; int seek = 0;
readBinaryFile(fname_weights, num, &mask_h, &mask_d, seek); readBinaryFile(fname_weights, num, &mask_h, &mask_d, seek);
seek += num; seek += num;
readBinaryFile(fname_weights, 3*num*2, &bias_h, &bias_d, seek); readBinaryFile(fname_weights, 3*num*2, &bias_h, &bias_d, seek);
} }
// init default classes name
classesNames.clear();
for(int i=0; i<classes; i++) {
classesNames.push_back(std::to_string(i));
}
// same // same
output_dim.n = input_dim.n; output_dim.n = input_dim.n;
output_dim.c = input_dim.c; output_dim.c = input_dim.c;
src/Yolo3Detection.cpp
+10 -7
View File
@@ -32,12 +32,13 @@ bool Yolo3Detection::init(std::string tensor_path) {
num = yRT->num; num = yRT->num;
// make a yolo layer for interpret predictions // make a yolo layer for interpret predictions
yolo[i] = new tk::dnn::Yolo(nullptr, classes, num, nullptr); // yolo without input and bias yolo[i] = new tk::dnn::Yolo(nullptr, classes, num, ""); // yolo without input and bias
yolo[i]->mask_h = new dnnType[num]; yolo[i]->mask_h = new dnnType[num];
yolo[i]->bias_h = new dnnType[num*3*2]; yolo[i]->bias_h = new dnnType[num*3*2];
memcpy(yolo[i]->mask_h, yRT->mask, sizeof(dnnType)*num); memcpy(yolo[i]->mask_h, yRT->mask, sizeof(dnnType)*num);
memcpy(yolo[i]->bias_h, yRT->bias, sizeof(dnnType)*num*3*2); memcpy(yolo[i]->bias_h, yRT->bias, sizeof(dnnType)*num*3*2);
yolo[i]->input_dim = yolo[i]->output_dim = tk::dnn::dataDim_t(1, yRT->c, yRT->h, yRT->w); yolo[i]->input_dim = yolo[i]->output_dim = tk::dnn::dataDim_t(1, yRT->c, yRT->h, yRT->w);
yolo[i]->classesNames = yRT->classesNames;
} }
dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes); dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes);
@@ -137,12 +138,12 @@ void Yolo3Detection::update(cv::Mat &imageORIG) {
cv::split(imageF,bgr);//split source cv::split(imageF,bgr);//split source
//write channels //write channels
int idx = 0; for(int i=0; i<netRT->input_dim.c; i++) {
memcpy((void*)&input[idx], (void*)bgr[2].data, imageF.rows*imageF.cols*sizeof(dnnType)); int idx = i*imageF.rows*imageF.cols;
idx = imageF.rows*imageF.cols; int ch = netRT->input_dim.c-1 -i;
memcpy((void*)&input[idx], (void*)bgr[1].data, imageF.rows*imageF.cols*sizeof(dnnType)); memcpy((void*)&input[idx], (void*)bgr[ch].data, imageF.rows*imageF.cols*sizeof(dnnType));
idx *= 2; }
memcpy((void*)&input[idx], (void*)bgr[0].data, imageF.rows*imageF.cols*sizeof(dnnType));
//DO INFERENCE //DO INFERENCE
dnnType *rt_out[3]; dnnType *rt_out[3];
@@ -155,6 +156,8 @@ void Yolo3Detection::update(cv::Mat &imageORIG) {
netRT->infer(dim, input_d); netRT->infer(dim, input_d);
TIMER_STOP TIMER_STOP
dim.print(); dim.print();
stats.push_back(t_ns);
} }
TIMER_START TIMER_START
src/utils.cpp
+5 -4
View File
@@ -21,7 +21,7 @@ bool fileExist(const char *fname) {
} }
void readBinaryFile(const char* fname, int size, dnnType** data_h, dnnType** data_d, int seek) 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::ifstream dataFile (fname, std::ios::in | std::ios::binary);
std::stringstream error_s; std::stringstream error_s;
@@ -39,7 +39,8 @@ void readBinaryFile(const char* fname, int size, dnnType** data_h, dnnType** dat
*data_h = new dnnType[size]; *data_h = new dnnType[size];
if (!dataFile.read ((char*) *data_h, size_b)) if (!dataFile.read ((char*) *data_h, size_b))
{ {
error_s << "Error reading file " << fname; error_s << "Error reading file " << fname << " with n of float: "<<size;
error_s << " seek: "<<seek << " size: "<<size_b<<"\n";
FatalError(error_s.str()); FatalError(error_s.str());
} }
@@ -67,7 +68,7 @@ void printDeviceVector(int size, dnnType* vec_d, bool device)
delete [] vec; 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; dnnType *data_h, *correct_h;
const float eps = 0.02f; const float eps = 0.02f;
@@ -91,7 +92,7 @@ int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device) {
diffs += 1; diffs += 1;
if(diffs == 1) if(diffs == 1)
std::cout<<"\n"; std::cout<<"\n";
if(diffs < 10) if(diffs < limit)
std::cout<<" | [ "<<i<<" ]: "<<data_h[i]<<" "<<correct_h[i]<<"\n"; std::cout<<" | [ "<<i<<" ]: "<<data_h[i]<<" "<<correct_h[i]<<"\n";
} }
} }
tests/imuodom/imuodom.cpp
+83
View File
@@ -0,0 +1,83 @@
#include<iostream>
#include "tkdnn.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() {
// 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(), &i0_h, &i0_d);
readBinaryFile(i1_bin, dim1.tot(), &i1_h, &i1_d);
readBinaryFile(i2_bin, dim2.tot(), &i2_h, &i2_d);
tk::dnn::Network net(dim0);
tk::dnn::Input x0 (&net, dim0, i0_d);
tk::dnn::Conv2d x0_0(&net, 128, 1, 11, 1, 1, 0, 0, c0_bin);
tk::dnn::Conv2d x0_1(&net, 128, 1, 11, 1, 1, 0, 0, c1_bin);
tk::dnn::Pooling x0_2(&net, 1, 3, 1, 3, tk::dnn::tkdnnPoolingMode_t::POOLING_MAX);
tk::dnn::Input x1 (&net, dim1, i1_d);
tk::dnn::Conv2d x1_0(&net, 128, 1, 11, 1, 1, 0, 0, c2_bin);
tk::dnn::Conv2d x1_1(&net, 128, 1, 11, 1, 1, 0, 0, c3_bin);
tk::dnn::Pooling x1_2(&net, 1, 3, 1, 3, tk::dnn::tkdnnPoolingMode_t::POOLING_MAX);
tk::dnn::Input x2 (&net, dim2, i2_d);
tk::dnn::Conv2d x2_0(&net, 128, 1, 11, 1, 1, 0, 0, c4_bin);
tk::dnn::Conv2d x2_1(&net, 128, 1, 11, 1, 1, 0, 0, c5_bin);
tk::dnn::Pooling x2_2(&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 (&net, concat_l, 3);
tk::dnn::LSTM lstm0(&net, 128, true, l0_bin);
tk::dnn::LSTM lstm1(&net, 128, false, l1_bin);
tk::dnn::Dense d0 (&net, 3, d0_bin);
tk::dnn::Layer *lstm1_l[1] = { &lstm1 };
tk::dnn::Route lstm1_link (&net, lstm1_l, 1);
tk::dnn::Dense d1 (&net, 4, d1_bin);
net.print();
dnnType *data;
tk::dnn::dataDim_t dim;
TIMER_START
// Inference
data = net.infer(dim, data);
TIMER_STOP
// Print real test
std::cout<<"\n==== CHECK RESULT ====\n";
dnnType *out0, *out1;
dnnType *out0_h, *out1_h;
readBinaryFile(o0_bin, d0.output_dim.tot(), &out0_h, &out0);
readBinaryFile(o1_bin, d1.output_dim.tot(), &out1_h, &out1);
d0.output_dim.print();
checkResult(d0.output_dim.tot(), d0.dstData, out0);
d1.output_dim.print();
checkResult(d1.output_dim.tot(), d1.dstData, out1);
return 0;
}
tests/imuodom/infer.py
+78
View File
@@ -0,0 +1,78 @@
import keras
from keras.models import load_model
import keras.backend.tensorflow_backend as KTF
import numpy as np
import argparse
import tensorflow as tf
import os
import random
import struct
from keras.models import Sequential, Model
def bin_write(f, data):
data = data.flatten()
fmt = 'f'*len(data)
bin = struct.pack(fmt, *data)
f.write(bin)
if __name__ == '__main__':
print("DATA FORMAT: ", keras.backend.image_data_format())
print("Load model: ", "ferrariS1.hdf5")
model = load_model("ferrariS1.hdf5")
model.summary()
weights = model.get_weights()
np.random.seed(2)
x_angle = np.random.rand(1,100,4)
x_gyro = np.random.rand(1,100,3)
x_acc = np.random.rand(1,100,3)
[yhat_delta_p, yhat_delta_q] = model.predict([x_angle, x_gyro, x_acc], batch_size=1, verbose=1)
#layer_name = 'dense_4'
#intermediate_layer_model = Model(inputs=model.input,
# outputs=model.get_layer(layer_name).output)
#intermediate_output = intermediate_layer_model.predict([x_angle, x_gyro, x_acc])
x_angle = np.array([x_angle])
x_gyro = np.array([x_gyro])
x_acc = np.array([x_acc])
#intermediate_output = np.array([intermediate_output])
x_angle = x_angle.transpose(0, 3, 1, 2)
x_gyro = x_gyro.transpose(0, 3, 1, 2)
x_acc = x_acc.transpose(0, 3, 1, 2)
#intermediate_output = intermediate_output.transpose(0, 3, 1, 2)
#print("Aggregate:")
#print(intermediate_output.tolist())
print("x0: ", np.shape(x_angle))
#print("out: ",np.shape(intermediate_output))
x_angle = np.array(x_angle.flatten(), dtype=np.float32)
x_gyro = np.array(x_gyro.flatten(), dtype=np.float32)
x_acc = np.array(x_acc.flatten(), dtype=np.float32)
yhat_delta_p = np.array(yhat_delta_p.flatten(), dtype=np.float32)
yhat_delta_q = np.array(yhat_delta_q.flatten(), dtype=np.float32)
#intermediate_output = np.array(intermediate_output.flatten(), dtype=np.float32)
f = open("layers/input0.bin", mode='wb')
bin_write(f, x_angle)
f = open("layers/input1.bin", mode='wb')
bin_write(f, x_gyro)
f = open("layers/input2.bin", mode='wb')
bin_write(f, x_acc)
f = open("layers/output0.bin", mode='wb')
bin_write(f, yhat_delta_p)
f = open("layers/output1.bin", mode='wb')
bin_write(f, yhat_delta_q)
#f = open("layers/output.bin", mode='wb')
#bin_write(f, intermediate_output)
tests/simple/test_model.py
+38 -27
View File
@@ -1,43 +1,54 @@
import keras import keras
import numpy as np import numpy as np
from keras.models import Sequential from keras.models import Sequential
from keras.layers import Input, Dense, Activation, Flatten, Dropout, ELU, Reshape, Lambda from keras.layers import Input, Dense, Activation, Flatten, Dropout, ELU, Reshape, Lambda, Conv1D
from keras.layers import Bidirectional, CuDNNLSTM
from keras.layers.convolutional import Convolution2D, Convolution3D from keras.layers.convolutional import Convolution2D, Convolution3D
from keras.layers.pooling import MaxPooling2D, MaxPooling3D, AveragePooling3D from keras.layers.pooling import MaxPooling2D, MaxPooling3D, AveragePooling3D
from keras.models import Sequential, Model from keras.models import Sequential, Model
from keras.layers import Cropping2D from keras.layers import Cropping2D
import keras.backend.tensorflow_backend as KTF import keras.backend.tensorflow_backend as KTF
import struct
from keras.models import Sequential, Model
def dense_model(): def bin_write(f, data):
model = Sequential() data = data.flatten()
fmt = 'f'*len(data)
bin = struct.pack(fmt, *data)
f.write(bin)
def create_model():
x1 = Input((3, 8), name='x1')
conv = Conv1D(4, 2)(x1)
lstm = Bidirectional(CuDNNLSTM(5, return_sequences=True))(conv)
lstm2 = Bidirectional(CuDNNLSTM(5, return_sequences=False))(lstm)
model = Model([x1], [lstm2])
model.summary()
model.add(Reshape((10, 10, 1), input_shape=(10, 10)))
model.add(Convolution2D(2, (4, 4), subsample=(2, 2),
bias_initializer='random_uniform', activation="relu"))
model.add(Convolution2D(4, (2, 2), subsample=(1, 1),
bias_initializer='random_uniform', activation="relu"))
model.add(Flatten())
model.add(Dense(4, bias_initializer='random_uniform', activation="relu"))
sgd = keras.optimizers.Adam(lr=1e-4, decay=1e-8)
model.compile(optimizer=sgd, loss="mse")
return model return model
if __name__ == '__main__': if __name__ == '__main__':
print "DATA FORMAT: ", keras.backend.image_data_format() print ("DATA FORMAT: ", keras.backend.image_data_format())
model = dense_model() model = create_model()
model.save("net.h5") model.save("net.hdf5")
grid = np.random.rand(10,10) np.random.seed(2)
X = grid[None,:,:] x = np.random.rand(1,1,3,8)
i = np.array(grid.flatten(), dtype=np.float32) r = model.predict( x[0], batch_size=1)
print i
i.tofile("input.bin", format="f") r = np.array([r])
print "Input: ", X x = x.transpose(0, 3, 1, 2)
#r = r.transpose(0, 3, 1, 2)
print("in: ", np.shape(x))
print("out: ", np.shape(r))
print("output: ", r.tolist())
x = np.array(x.flatten(), dtype=np.float32)
f = open("input.bin", mode='wb')
bin_write(f, x)
r = np.array(r.flatten(), dtype=np.float32)
f = open("output.bin", mode='wb')
bin_write(f, r)
r = model.predict( X, batch_size=1)
print np.shape(r)
print "Result: ", r
print "Result shape: ", np.shape(r)
r.tofile("output.bin", format="f")
tests/simple/test_simple.cpp
+9 -11
View File
@@ -2,23 +2,21 @@
#include "tkdnn.h" #include "tkdnn.h"
const char *input_bin = "../tests/simple/input.bin"; const char *input_bin = "../tests/simple/input.bin";
const char *c0_bin = "../tests/simple/layers/c0.bin"; const char *c0_bin = "../tests/simple/layers/conv1d_1.bin";
const char *c1_bin = "../tests/simple/layers/c1.bin"; const char *l1_bin = "../tests/simple/layers/bidirectional_1.bin";
const char *d2_bin = "../tests/simple/layers/d2.bin"; const char *l2_bin = "../tests/simple/layers/bidirectional_2.bin";
const char *output_bin = "../tests/simple/output.bin"; const char *output_bin = "../tests/simple/output.bin";
int main() { int main() {
// Network layout // Network layout
tk::dnn::dataDim_t dim(1, 1, 10, 10, 1); tk::dnn::dataDim_t dim(1, 8, 1, 3);
tk::dnn::Network net(dim); tk::dnn::Network net(dim);
tk::dnn::Conv2d l0(&net, 2, 4, 4, 2, 2, 0, 0, c0_bin); tk::dnn::Conv2d l0(&net, 4, 1, 2, 1, 1, 0, 0, c0_bin);
tk::dnn::Activation l1(&net, CUDNN_ACTIVATION_RELU); tk::dnn::LSTM l1(&net, 5, true, l1_bin);
tk::dnn::Conv2d l2(&net, 4, 2, 2, 1, 1, 0, 0, c1_bin); tk::dnn::LSTM l2(&net, 5, false, l2_bin);
tk::dnn::Activation l3(&net, CUDNN_ACTIVATION_RELU);
tk::dnn::Flatten l4(&net); net.print();
tk::dnn::Dense l5(&net, 4, d2_bin);
tk::dnn::Activation l6(&net, CUDNN_ACTIVATION_RELU);
// Load input // Load input
dnnType *data; dnnType *data;
tests/weights_exporter.py
+90 -85
View File
@@ -5,96 +5,89 @@ import numpy as np
import argparse import argparse
import tensorflow as tf import tensorflow as tf
import os import os
import msgpack
import lmdb
import random import random
import struct
from keras.models import Sequential, Model
def export_dense(name, weights, bias): def bin_write(f, data):
print "######## EXPORT", name, "LAYER ########" data = data.flatten()
print "Original weighs:" fmt = 'f'*len(data)
print weights bin = struct.pack(fmt, *data)
print bias, "\n" f.write(bin)
#input, filters def export_layer(name, weights, bias):
I, C = np.shape(weights) print ("######## EXPORT", name, "LAYER ########")
B = np.shape(bias)
print "w shape: ", I, C
print "b shape: ", B
wgs = [ [ j[i] for j in weights ] for i in xrange(C) ] print("wgs pretranpose: ", np.shape(weights))
wgs = np.array(wgs, dtype=np.float32) # convert NHWC to NCHW
if(weights.ndim == 4):
weights = weights.transpose(3,2,0,1)
elif(weights.ndim == 3):
weights = weights.transpose(2,1,0)
elif(weights.ndim == 2):
weights = weights.transpose(1,0)
else:
print("Ndim", weights.ndim)
raise("not implemented with dim" )
print "REPOSITIONED WEIGHTS:" print("weights: ", np.shape(weights))
print wgs print("bias: ", np.shape(bias))
weights = np.array(weights.flatten(), dtype=np.float32)
bias = np.array(bias, dtype=np.float32) bias = np.array(bias, dtype=np.float32)
wgs.tofile(name + ".bin", format="f") print(len(weights) + len(bias))
bias.tofile(name + ".bias.bin", format="f")
print "WEIGHTS saved\n"
def export_conv2d(name, weights, bias): f = open(name + ".bin", mode='wb')
print "######## EXPORT", name, "LAYER ########" bin_write(f, weights)
print "Original weighs:" bin_write(f, bias)
print weights print ("WEIGHTS saved\n")
print bias, "\n"
# height, width, input, filters def export_bidir(name, params, paramsb):
H, W, N, C = np.shape(weights) print ("######## EXPORT", name, "LAYER ########")
B = np.shape(bias)
print "w shape: ", N, C, H, W
print "b shape: ", B
wgs = weights.transpose() f = open(name + ".bin", mode='wb')
wgs = wgs.transpose(0, 1, 3, 2)
print "Final shape:", np.shape(wgs)
wgs = np.array(wgs.flatten(), dtype=np.float32)
print "REPOSITIONED WEIGHTS:" print("FORWARD")
print wgs ker = params[0]
rec_ker = params[1]
bias = np.array(bias, dtype=np.float32) bias = params[2]
print ("export kernels: ", np.shape(ker))
wgs.tofile(name + ".bin", format="f") units = np.shape(ker)[1] // 4
bias.tofile(name + ".bias.bin", format="f") bin_write(f, ker[:,:units])
print "WEIGHTS saved\n" bin_write(f, ker[:,units:units*2])
bin_write(f, ker[:,units*2:units*3])
def export_conv3d(name, weights, bias): bin_write(f, ker[:,units*3:])
print "######## EXPORT", name, "LAYER ########" print ("export recurrent kernels: ", np.shape(rec_ker))
print "Original weighs:" bin_write(f, rec_ker[:,:units])
print weights bin_write(f, rec_ker[:,units:units*2])
print bias, "\n" bin_write(f, rec_ker[:,units*2:units*3])
bin_write(f, rec_ker[:,units*3:])
print np.shape(weights) print ("export kernels: ", np.shape(ker))
# height, width, input, thickness, filters bin_write(f, bias)
H, W, T, N, C = np.shape(weights) print("WEIGHTS saved\n")
B = np.shape(bias)
print "w shape: ", T, C, H, W #thickness is number of images for cudnn
print "b shape: ", B
wgs = weights.transpose()
wgs = wgs.transpose(0, 1, 4, 3, 2)
print "Final shape:", np.shape(wgs)
wgs = np.array(wgs.flatten(), dtype=np.float32)
print "REPOSITIONED WEIGHTS:"
print wgs
bias = np.array(bias, dtype=np.float32)
wgs.tofile(name + ".bin", format="f")
bias.tofile(name + ".bias.bin", format="f")
print "WEIGHTS saved\n"
def get_session(gpu_fraction=0.5):
gpu_options = tf.GPUOptions(allow_growth=True)
#per_process_gpu_memory_fraction=gpu_fraction)
return tf.Session(config=tf.ConfigProto(gpu_options=gpu_options))
print("BACKWARD")
ker = paramsb[0]
rec_ker = paramsb[1]
bias = paramsb[2]
print ("export kernels: ", np.shape(ker))
units = np.shape(ker)[1] // 4
bin_write(f, ker[:,:units])
bin_write(f, ker[:,units:units*2])
bin_write(f, ker[:,units*2:units*3])
bin_write(f, ker[:,units*3:])
print ("export recurrent kernels: ", np.shape(rec_ker))
bin_write(f, rec_ker[:,:units])
bin_write(f, rec_ker[:,units:units*2])
bin_write(f, rec_ker[:,units*2:units*3])
bin_write(f, rec_ker[:,units*3:])
print ("export kernels: ", np.shape(ker))
bin_write(f, bias)
print("WEIGHTS saved\n")
#https://github.com/fchollet/keras/wiki/Converting-convolution-kernels-from-Theano-to-TensorFlow-and-vice-versa #https://github.com/fchollet/keras/wiki/Converting-convolution-kernels-from-Theano-to-TensorFlow-and-vice-versa
if __name__ == '__main__': if __name__ == '__main__':
KTF.set_session(get_session()) print("DATA FORMAT: ", keras.backend.image_data_format())
parser = argparse.ArgumentParser(description='KERAS WEIGHTS EXPORTER TO CUDNN') parser = argparse.ArgumentParser(description='KERAS WEIGHTS EXPORTER TO CUDNN')
parser.add_argument('model',type=str, parser.add_argument('model',type=str,
@@ -103,31 +96,43 @@ if __name__ == '__main__':
args = parser.parse_args() args = parser.parse_args()
print "DATA FORMAT: ", keras.backend.image_data_format() print("DATA FORMAT: ", keras.backend.image_data_format())
print "Load model: ", args.model print("Load model: ", args.model)
model = load_model(args.model) model = load_model(args.model)
model.summary()
weights = model.get_weights() weights = model.get_weights()
ws = np.shape(weights) ws = np.shape(weights)
print "Weights shape:", ws print("Weights shape:", ws)
if not os.path.exists(args.output): if not os.path.exists(args.output):
os.makedirs(args.output) os.makedirs(args.output)
num = 0
name_num = 0 name_num = 0
for l in model.layers: for l in model.layers:
print("\n\nNAME: ", l.name)
print("input: ", l.input_shape, " output: ", l.output_shape)
wgs = l.get_weights()
print("wgs num: ", len(wgs))
name = l.name name = l.name
if name.startswith("conv3d"): if name.startswith("conv3d"):
export_conv3d(args.output + "/conv" + str(name_num), weights[num], weights[num+1]) export_layer(args.output + "/" + name, wgs[0], wgs[1])
elif name.startswith("conv2d"): elif name.startswith("conv2d"):
export_conv2d(args.output + "/conv" + str(name_num), weights[num], weights[num+1]) export_layer(args.output + "/" + name, wgs[0], wgs[1])
elif name.startswith("conv1d"):
export_layer(args.output + "/" + name, wgs[0], wgs[1])
elif name.startswith("dense"): elif name.startswith("dense"):
export_dense(args.output + "/dense" + str(name_num), weights[num], weights[num+1]) export_layer(args.output + "/" + name, wgs[0], wgs[1])
elif name.startswith("bidirectional"):
wgs = l.forward_layer.get_weights()
export_bidir(args.output + "/" + name, l.forward_layer.get_weights(), l.backward_layer.get_weights())
else: else:
print "skip:", name, "has no weights" print ("skip:", name, "has no weights")
continue continue
name_num += 1
num += 2
tests/yolo3_berkeley/yolo3_berkeley.cpp
+14 -295
View File
@@ -1,295 +1,23 @@
#include<iostream> #include<iostream>
#include<vector>
#include "tkdnn.h" #include "tkdnn.h"
const char *input_bin = "../tests/yolo3_berkeley/layers/input.bin";
const char *c0_bin = "../tests/yolo3_berkeley/layers/c0.bin";
const char *c1_bin = "../tests/yolo3_berkeley/layers/c1.bin";
const char *c2_bin = "../tests/yolo3_berkeley/layers/c2.bin";
const char *c3_bin = "../tests/yolo3_berkeley/layers/c3.bin";
const char *c5_bin = "../tests/yolo3_berkeley/layers/c5.bin";
const char *c6_bin = "../tests/yolo3_berkeley/layers/c6.bin";
const char *c7_bin = "../tests/yolo3_berkeley/layers/c7.bin";
const char *c9_bin = "../tests/yolo3_berkeley/layers/c9.bin";
const char *c10_bin = "../tests/yolo3_berkeley/layers/c10.bin";
const char *c12_bin = "../tests/yolo3_berkeley/layers/c12.bin";
const char *c13_bin = "../tests/yolo3_berkeley/layers/c13.bin";
const char *c14_bin = "../tests/yolo3_berkeley/layers/c14.bin";
const char *c16_bin = "../tests/yolo3_berkeley/layers/c16.bin";
const char *c17_bin = "../tests/yolo3_berkeley/layers/c17.bin";
const char *c19_bin = "../tests/yolo3_berkeley/layers/c19.bin";
const char *c20_bin = "../tests/yolo3_berkeley/layers/c20.bin";
const char *c22_bin = "../tests/yolo3_berkeley/layers/c22.bin";
const char *c23_bin = "../tests/yolo3_berkeley/layers/c23.bin";
const char *c25_bin = "../tests/yolo3_berkeley/layers/c25.bin";
const char *c26_bin = "../tests/yolo3_berkeley/layers/c26.bin";
const char *c28_bin = "../tests/yolo3_berkeley/layers/c28.bin";
const char *c29_bin = "../tests/yolo3_berkeley/layers/c29.bin";
const char *c31_bin = "../tests/yolo3_berkeley/layers/c31.bin";
const char *c32_bin = "../tests/yolo3_berkeley/layers/c32.bin";
const char *c34_bin = "../tests/yolo3_berkeley/layers/c34.bin";
const char *c35_bin = "../tests/yolo3_berkeley/layers/c35.bin";
const char *c37_bin = "../tests/yolo3_berkeley/layers/c37.bin";
const char *c38_bin = "../tests/yolo3_berkeley/layers/c38.bin";
const char *c39_bin = "../tests/yolo3_berkeley/layers/c39.bin";
const char *c41_bin = "../tests/yolo3_berkeley/layers/c41.bin";
const char *c42_bin = "../tests/yolo3_berkeley/layers/c42.bin";
const char *c44_bin = "../tests/yolo3_berkeley/layers/c44.bin";
const char *c45_bin = "../tests/yolo3_berkeley/layers/c45.bin";
const char *c47_bin = "../tests/yolo3_berkeley/layers/c47.bin";
const char *c48_bin = "../tests/yolo3_berkeley/layers/c48.bin";
const char *c50_bin = "../tests/yolo3_berkeley/layers/c50.bin";
const char *c51_bin = "../tests/yolo3_berkeley/layers/c51.bin";
const char *c53_bin = "../tests/yolo3_berkeley/layers/c53.bin";
const char *c54_bin = "../tests/yolo3_berkeley/layers/c54.bin";
const char *c56_bin = "../tests/yolo3_berkeley/layers/c56.bin";
const char *c57_bin = "../tests/yolo3_berkeley/layers/c57.bin";
const char *c59_bin = "../tests/yolo3_berkeley/layers/c59.bin";
const char *c60_bin = "../tests/yolo3_berkeley/layers/c60.bin";
const char *c62_bin = "../tests/yolo3_berkeley/layers/c62.bin";
const char *c63_bin = "../tests/yolo3_berkeley/layers/c63.bin";
const char *c64_bin = "../tests/yolo3_berkeley/layers/c64.bin";
const char *c66_bin = "../tests/yolo3_berkeley/layers/c66.bin";
const char *c67_bin = "../tests/yolo3_berkeley/layers/c67.bin";
const char *c69_bin = "../tests/yolo3_berkeley/layers/c69.bin";
const char *c70_bin = "../tests/yolo3_berkeley/layers/c70.bin";
const char *c72_bin = "../tests/yolo3_berkeley/layers/c72.bin";
const char *c73_bin = "../tests/yolo3_berkeley/layers/c73.bin";
const char *c75_bin = "../tests/yolo3_berkeley/layers/c75.bin";
const char *c76_bin = "../tests/yolo3_berkeley/layers/c76.bin";
const char *c77_bin = "../tests/yolo3_berkeley/layers/c77.bin";
const char *c78_bin = "../tests/yolo3_berkeley/layers/c78.bin";
const char *c79_bin = "../tests/yolo3_berkeley/layers/c79.bin";
const char *c80_bin = "../tests/yolo3_berkeley/layers/c80.bin";
const char *c81_bin = "../tests/yolo3_berkeley/layers/c81.bin";
const char *g82_bin = "../tests/yolo3_berkeley/layers/g82.bin";
const char *c84_bin = "../tests/yolo3_berkeley/layers/c84.bin";
const char *c87_bin = "../tests/yolo3_berkeley/layers/c87.bin";
const char *c88_bin = "../tests/yolo3_berkeley/layers/c88.bin";
const char *c89_bin = "../tests/yolo3_berkeley/layers/c89.bin";
const char *c90_bin = "../tests/yolo3_berkeley/layers/c90.bin";
const char *c91_bin = "../tests/yolo3_berkeley/layers/c91.bin";
const char *c92_bin = "../tests/yolo3_berkeley/layers/c92.bin";
const char *c93_bin = "../tests/yolo3_berkeley/layers/c93.bin";
const char *g94_bin = "../tests/yolo3_berkeley/layers/g94.bin";
const char *c96_bin = "../tests/yolo3_berkeley/layers/c96.bin";
const char *c99_bin = "../tests/yolo3_berkeley/layers/c99.bin";
const char *c100_bin = "../tests/yolo3_berkeley/layers/c100.bin";
const char *c101_bin = "../tests/yolo3_berkeley/layers/c101.bin";
const char *c102_bin = "../tests/yolo3_berkeley/layers/c102.bin";
const char *c103_bin = "../tests/yolo3_berkeley/layers/c103.bin";
const char *c104_bin = "../tests/yolo3_berkeley/layers/c104.bin";
const char *c105_bin = "../tests/yolo3_berkeley/layers/c105.bin";
const char *g106_bin = "../tests/yolo3_berkeley/layers/g106.bin";
const char *output_bins[3] = {
"../tests/yolo3_berkeley/debug/layer82_out.bin",
"../tests/yolo3_berkeley/debug/layer94_out.bin",
"../tests/yolo3_berkeley/debug/layer106_out.bin"
};
int main() { int main() {
// Network layout // Network layout
tk::dnn::dataDim_t dim(1, 3, 320, 544, 1); tk::dnn::dataDim_t dim(1, 3, 320, 544, 1);
tk::dnn::Network net(dim); tk::dnn::Network net(dim);
tk::dnn::Conv2d c0 (&net, 32, 3, 3, 1, 1, 1, 1, c0_bin, true); // create yolo3 model
tk::dnn::Activation a0 (&net, tk::dnn::ACTIVATION_LEAKY); std::string bin_path = "../tests/yolo3_berkeley";
tk::dnn::Conv2d c1 (&net, 64, 3, 3, 2, 2, 1, 1, c1_bin, true); int classes = 10;
tk::dnn::Activation a1 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Yolo *yolo [3];
tk::dnn::Conv2d c2 (&net, 32, 1, 1, 1, 1, 0, 0, c2_bin, true); #include "models/Yolo3.h"
tk::dnn::Activation a2 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c3 (&net, 64, 3, 3, 1, 1, 1, 1, c3_bin, true);
tk::dnn::Activation a3 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s4 (&net, &a1);
tk::dnn::Conv2d c5 (&net, 128, 3, 3, 2, 2, 1, 1, c5_bin, true);
tk::dnn::Activation a5 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c6 (&net, 64, 1, 1, 1, 1, 0, 0, c6_bin, true);
tk::dnn::Activation a6 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c7 (&net, 128, 3, 3, 1, 1, 1, 1, c7_bin, true);
tk::dnn::Activation a7 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s8 (&net, &a5);
tk::dnn::Conv2d c9 (&net, 64, 1, 1, 1, 1, 0, 0, c9_bin, true);
tk::dnn::Activation a9 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c10 (&net, 128, 3, 3, 1, 1, 1, 1, c10_bin, true);
tk::dnn::Activation a10 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s11 (&net, &s8);
tk::dnn::Conv2d c12 (&net, 256, 3, 3, 2, 2, 1, 1, c12_bin, true); // fill classes names
tk::dnn::Activation a12 (&net, tk::dnn::ACTIVATION_LEAKY); for(int i=0; i<3; i++) {
tk::dnn::Conv2d c13 (&net, 128, 1, 1, 1, 1, 0, 0, c13_bin, true); yolo[i]->classesNames = {"person", "car", "truck", "bus", "motor", "bike", "rider", "traffic light", "traffic sign", "train"};
tk::dnn::Activation a13 (&net, tk::dnn::ACTIVATION_LEAKY); }
tk::dnn::Conv2d c14 (&net, 256, 3, 3, 1, 1, 1, 1, c14_bin, true);
tk::dnn::Activation a14 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s15 (&net, &a12);
tk::dnn::Conv2d c16 (&net, 128, 1, 1, 1, 1, 0, 0, c16_bin, true);
tk::dnn::Activation a16 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c17 (&net, 256, 3, 3, 1, 1, 1, 1, c17_bin, true);
tk::dnn::Activation a17 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s18 (&net, &s15);
tk::dnn::Conv2d c19 (&net, 128, 1, 1, 1, 1, 0, 0, c19_bin, true);
tk::dnn::Activation a19 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c20 (&net, 256, 3, 3, 1, 1, 1, 1, c20_bin, true);
tk::dnn::Activation a20 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s21 (&net, &s18);
tk::dnn::Conv2d c22 (&net, 128, 1, 1, 1, 1, 0, 0, c22_bin, true);
tk::dnn::Activation a22 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c23 (&net, 256, 3, 3, 1, 1, 1, 1, c23_bin, true);
tk::dnn::Activation a23 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s24 (&net, &s21);
tk::dnn::Conv2d c25 (&net, 128, 1, 1, 1, 1, 0, 0, c25_bin, true);
tk::dnn::Activation a25 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c26 (&net, 256, 3, 3, 1, 1, 1, 1, c26_bin, true);
tk::dnn::Activation a26 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s27 (&net, &s24);
tk::dnn::Conv2d c28 (&net, 128, 1, 1, 1, 1, 0, 0, c28_bin, true);
tk::dnn::Activation a28 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c29 (&net, 256, 3, 3, 1, 1, 1, 1, c29_bin, true);
tk::dnn::Activation a29 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s30 (&net, &s27);
tk::dnn::Conv2d c31 (&net, 128, 1, 1, 1, 1, 0, 0, c31_bin, true);
tk::dnn::Activation a31 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c32 (&net, 256, 3, 3, 1, 1, 1, 1, c32_bin, true);
tk::dnn::Activation a32 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s33 (&net, &s30);
tk::dnn::Conv2d c34 (&net, 128, 1, 1, 1, 1, 0, 0, c34_bin, true);
tk::dnn::Activation a34 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c35 (&net, 256, 3, 3, 1, 1, 1, 1, c35_bin, true);
tk::dnn::Activation a35 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s36 (&net, &s33);
tk::dnn::Conv2d c37 (&net, 512, 3, 3, 2, 2, 1, 1, c37_bin, true);
tk::dnn::Activation a37 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c38 (&net, 256, 1, 1, 1, 1, 0, 0, c38_bin, true);
tk::dnn::Activation a38 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c39 (&net, 512, 3, 3, 1, 1, 1, 1, c39_bin, true);
tk::dnn::Activation a39 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s40 (&net, &a37);
tk::dnn::Conv2d c41 (&net, 256, 1, 1, 1, 1, 0, 0, c41_bin, true);
tk::dnn::Activation a41 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c42 (&net, 512, 3, 3, 1, 1, 1, 1, c42_bin, true);
tk::dnn::Activation a42 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s43 (&net, &s40);
tk::dnn::Conv2d c44 (&net, 256, 1, 1, 1, 1, 0, 0, c44_bin, true);
tk::dnn::Activation a44 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c45 (&net, 512, 3, 3, 1, 1, 1, 1, c45_bin, true);
tk::dnn::Activation a45 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s46 (&net, &s43);
tk::dnn::Conv2d c47 (&net, 256, 1, 1, 1, 1, 0, 0, c47_bin, true);
tk::dnn::Activation a47 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c48 (&net, 512, 3, 3, 1, 1, 1, 1, c48_bin, true);
tk::dnn::Activation a48 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s49 (&net, &s46);
tk::dnn::Conv2d c50 (&net, 256, 1, 1, 1, 1, 0, 0, c50_bin, true);
tk::dnn::Activation a50 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c51 (&net, 512, 3, 3, 1, 1, 1, 1, c51_bin, true);
tk::dnn::Activation a51 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s52 (&net, &s49);
tk::dnn::Conv2d c53 (&net, 256, 1, 1, 1, 1, 0, 0, c53_bin, true);
tk::dnn::Activation a53 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c54 (&net, 512, 3, 3, 1, 1, 1, 1, c54_bin, true);
tk::dnn::Activation a54 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s55 (&net, &s52);
tk::dnn::Conv2d c56 (&net, 256, 1, 1, 1, 1, 0, 0, c56_bin, true);
tk::dnn::Activation a56 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c57 (&net, 512, 3, 3, 1, 1, 1, 1, c57_bin, true);
tk::dnn::Activation a57 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s58 (&net, &s55);
tk::dnn::Conv2d c59 (&net, 256, 1, 1, 1, 1, 0, 0, c59_bin, true);
tk::dnn::Activation a59 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c60 (&net, 512, 3, 3, 1, 1, 1, 1, c60_bin, true);
tk::dnn::Activation a60 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s61 (&net, &s58);
tk::dnn::Conv2d c62 (&net,1024, 3, 3, 2, 2, 1, 1, c62_bin, true);
tk::dnn::Activation a62 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c63 (&net, 512, 1, 1, 1, 1, 0, 0, c63_bin, true);
tk::dnn::Activation a63 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c64 (&net,1024, 3, 3, 1, 1, 1, 1, c64_bin, true);
tk::dnn::Activation a64 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s65 (&net, &a62);
tk::dnn::Conv2d c66 (&net, 512, 1, 1, 1, 1, 0, 0, c66_bin, true);
tk::dnn::Activation a66 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c67 (&net,1024, 3, 3, 1, 1, 1, 1, c67_bin, true);
tk::dnn::Activation a67 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s68 (&net, &s65);
tk::dnn::Conv2d c69 (&net, 512, 1, 1, 1, 1, 0, 0, c69_bin, true);
tk::dnn::Activation a69 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c70 (&net,1024, 3, 3, 1, 1, 1, 1, c70_bin, true);
tk::dnn::Activation a70 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s71 (&net, &s68);
tk::dnn::Conv2d c72 (&net, 512, 1, 1, 1, 1, 0, 0, c72_bin, true);
tk::dnn::Activation a72 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c73 (&net,1024, 3, 3, 1, 1, 1, 1, c73_bin, true);
tk::dnn::Activation a73 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s74 (&net, &s71);
tk::dnn::Conv2d c75 (&net, 512, 1, 1, 1, 1, 0, 0, c75_bin, true);
tk::dnn::Activation a75 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c76 (&net,1024, 3, 3, 1, 1, 1, 1, c76_bin, true);
tk::dnn::Activation a76 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c77 (&net, 512, 1, 1, 1, 1, 0, 0, c77_bin, true);
tk::dnn::Activation a77 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c78 (&net,1024, 3, 3, 1, 1, 1, 1, c78_bin, true);
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);
tk::dnn::Yolo yolo0 (&net, 10, 3, g82_bin);
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::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);
tk::dnn::Activation a88 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c89 (&net, 256, 1, 1, 1, 1, 0, 0, c89_bin, true);
tk::dnn::Activation a89 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c90 (&net, 512, 3, 3, 1, 1, 1, 1, c90_bin, true);
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);
tk::dnn::Yolo yolo1 (&net, 10, 3, g94_bin);
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::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);
tk::dnn::Activation a100 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c101 (&net, 128, 1, 1, 1, 1, 0, 0, c101_bin, true);
tk::dnn::Activation a101 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c102 (&net, 256, 3, 3, 1, 1, 1, 1, c102_bin, true);
tk::dnn::Activation a102 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c103 (&net, 128, 1, 1, 1, 1, 0, 0, c103_bin, true);
tk::dnn::Activation a103 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c104 (&net, 256, 3, 3, 1, 1, 1, 1, c104_bin, true);
tk::dnn::Activation a104 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c105 (&net, 45, 1, 1, 1, 1, 0, 0, c105_bin, false);
tk::dnn::Yolo yolo2 (&net, 10, 3, g106_bin);
// Load input // Load input
dnnType *data; dnnType *data;
@@ -304,9 +32,7 @@ int main() {
// the network have 3 outputs // the network have 3 outputs
tk::dnn::dataDim_t out_dim[3]; tk::dnn::dataDim_t out_dim[3];
out_dim[0] = yolo0.output_dim; for(int i=0; i<3; i++) out_dim[i] = yolo[i]->output_dim;
out_dim[1] = yolo1.output_dim;
out_dim[2] = yolo2.output_dim;
dnnType *cudnn_out[3], *rt_out[3]; dnnType *cudnn_out[3], *rt_out[3];
tk::dnn::dataDim_t dim1 = dim; //input dim tk::dnn::dataDim_t dim1 = dim; //input dim
@@ -317,18 +43,13 @@ int main() {
TIMER_STOP TIMER_STOP
dim1.print(); dim1.print();
} }
cudnn_out[0] = yolo0.dstData; for(int i=0; i<3; i++) cudnn_out[i] = yolo[i]->dstData;
cudnn_out[1] = yolo1.dstData;
cudnn_out[2] = yolo2.dstData;
printCenteredTitle(" compute detections ", '=', 30); printCenteredTitle(" compute detections ", '=', 30);
TIMER_START TIMER_START
int ndets = 0; int ndets = 0;
int classes = yolo0.classes;
tk::dnn::Yolo::detection *dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes); tk::dnn::Yolo::detection *dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes);
yolo0.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5); for(int i=0; i<3; i++) yolo[i]->computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
yolo1.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
yolo2.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
tk::dnn::Yolo::mergeDetections(dets, ndets, classes); tk::dnn::Yolo::mergeDetections(dets, ndets, classes);
for(int j=0; j<ndets; j++) { for(int j=0; j<ndets; j++) {
@@ -356,9 +77,7 @@ int main() {
TIMER_STOP TIMER_STOP
dim2.print(); dim2.print();
} }
rt_out[0] = (dnnType*)netRT.buffersRT[1]; for(int i=0; i<3; i++) rt_out[i] = (dnnType*)netRT.buffersRT[i+1];
rt_out[1] = (dnnType*)netRT.buffersRT[2];
rt_out[2] = (dnnType*)netRT.buffersRT[3];
for(int i=0; i<3; i++) { for(int i=0; i<3; i++) {
printCenteredTitle((std::string(" YOLO ") + std::to_string(i) + " CHECK RESULTS ").c_str(), '=', 30); printCenteredTitle((std::string(" YOLO ") + std::to_string(i) + " CHECK RESULTS ").c_str(), '=', 30);
tests/yolo3_coco4/yolo3_coco4.cpp
+10 -296
View File
@@ -1,295 +1,18 @@
#include<iostream> #include<iostream>
#include<vector>
#include "tkdnn.h" #include "tkdnn.h"
const char *input_bin = "../tests/yolo3_coco4/layers/input.bin";
const char *c0_bin = "../tests/yolo3_coco4/layers/c0.bin";
const char *c1_bin = "../tests/yolo3_coco4/layers/c1.bin";
const char *c2_bin = "../tests/yolo3_coco4/layers/c2.bin";
const char *c3_bin = "../tests/yolo3_coco4/layers/c3.bin";
const char *c5_bin = "../tests/yolo3_coco4/layers/c5.bin";
const char *c6_bin = "../tests/yolo3_coco4/layers/c6.bin";
const char *c7_bin = "../tests/yolo3_coco4/layers/c7.bin";
const char *c9_bin = "../tests/yolo3_coco4/layers/c9.bin";
const char *c10_bin = "../tests/yolo3_coco4/layers/c10.bin";
const char *c12_bin = "../tests/yolo3_coco4/layers/c12.bin";
const char *c13_bin = "../tests/yolo3_coco4/layers/c13.bin";
const char *c14_bin = "../tests/yolo3_coco4/layers/c14.bin";
const char *c16_bin = "../tests/yolo3_coco4/layers/c16.bin";
const char *c17_bin = "../tests/yolo3_coco4/layers/c17.bin";
const char *c19_bin = "../tests/yolo3_coco4/layers/c19.bin";
const char *c20_bin = "../tests/yolo3_coco4/layers/c20.bin";
const char *c22_bin = "../tests/yolo3_coco4/layers/c22.bin";
const char *c23_bin = "../tests/yolo3_coco4/layers/c23.bin";
const char *c25_bin = "../tests/yolo3_coco4/layers/c25.bin";
const char *c26_bin = "../tests/yolo3_coco4/layers/c26.bin";
const char *c28_bin = "../tests/yolo3_coco4/layers/c28.bin";
const char *c29_bin = "../tests/yolo3_coco4/layers/c29.bin";
const char *c31_bin = "../tests/yolo3_coco4/layers/c31.bin";
const char *c32_bin = "../tests/yolo3_coco4/layers/c32.bin";
const char *c34_bin = "../tests/yolo3_coco4/layers/c34.bin";
const char *c35_bin = "../tests/yolo3_coco4/layers/c35.bin";
const char *c37_bin = "../tests/yolo3_coco4/layers/c37.bin";
const char *c38_bin = "../tests/yolo3_coco4/layers/c38.bin";
const char *c39_bin = "../tests/yolo3_coco4/layers/c39.bin";
const char *c41_bin = "../tests/yolo3_coco4/layers/c41.bin";
const char *c42_bin = "../tests/yolo3_coco4/layers/c42.bin";
const char *c44_bin = "../tests/yolo3_coco4/layers/c44.bin";
const char *c45_bin = "../tests/yolo3_coco4/layers/c45.bin";
const char *c47_bin = "../tests/yolo3_coco4/layers/c47.bin";
const char *c48_bin = "../tests/yolo3_coco4/layers/c48.bin";
const char *c50_bin = "../tests/yolo3_coco4/layers/c50.bin";
const char *c51_bin = "../tests/yolo3_coco4/layers/c51.bin";
const char *c53_bin = "../tests/yolo3_coco4/layers/c53.bin";
const char *c54_bin = "../tests/yolo3_coco4/layers/c54.bin";
const char *c56_bin = "../tests/yolo3_coco4/layers/c56.bin";
const char *c57_bin = "../tests/yolo3_coco4/layers/c57.bin";
const char *c59_bin = "../tests/yolo3_coco4/layers/c59.bin";
const char *c60_bin = "../tests/yolo3_coco4/layers/c60.bin";
const char *c62_bin = "../tests/yolo3_coco4/layers/c62.bin";
const char *c63_bin = "../tests/yolo3_coco4/layers/c63.bin";
const char *c64_bin = "../tests/yolo3_coco4/layers/c64.bin";
const char *c66_bin = "../tests/yolo3_coco4/layers/c66.bin";
const char *c67_bin = "../tests/yolo3_coco4/layers/c67.bin";
const char *c69_bin = "../tests/yolo3_coco4/layers/c69.bin";
const char *c70_bin = "../tests/yolo3_coco4/layers/c70.bin";
const char *c72_bin = "../tests/yolo3_coco4/layers/c72.bin";
const char *c73_bin = "../tests/yolo3_coco4/layers/c73.bin";
const char *c75_bin = "../tests/yolo3_coco4/layers/c75.bin";
const char *c76_bin = "../tests/yolo3_coco4/layers/c76.bin";
const char *c77_bin = "../tests/yolo3_coco4/layers/c77.bin";
const char *c78_bin = "../tests/yolo3_coco4/layers/c78.bin";
const char *c79_bin = "../tests/yolo3_coco4/layers/c79.bin";
const char *c80_bin = "../tests/yolo3_coco4/layers/c80.bin";
const char *c81_bin = "../tests/yolo3_coco4/layers/c81.bin";
const char *g82_bin = "../tests/yolo3_coco4/layers/g82.bin";
const char *c84_bin = "../tests/yolo3_coco4/layers/c84.bin";
const char *c87_bin = "../tests/yolo3_coco4/layers/c87.bin";
const char *c88_bin = "../tests/yolo3_coco4/layers/c88.bin";
const char *c89_bin = "../tests/yolo3_coco4/layers/c89.bin";
const char *c90_bin = "../tests/yolo3_coco4/layers/c90.bin";
const char *c91_bin = "../tests/yolo3_coco4/layers/c91.bin";
const char *c92_bin = "../tests/yolo3_coco4/layers/c92.bin";
const char *c93_bin = "../tests/yolo3_coco4/layers/c93.bin";
const char *g94_bin = "../tests/yolo3_coco4/layers/g94.bin";
const char *c96_bin = "../tests/yolo3_coco4/layers/c96.bin";
const char *c99_bin = "../tests/yolo3_coco4/layers/c99.bin";
const char *c100_bin = "../tests/yolo3_coco4/layers/c100.bin";
const char *c101_bin = "../tests/yolo3_coco4/layers/c101.bin";
const char *c102_bin = "../tests/yolo3_coco4/layers/c102.bin";
const char *c103_bin = "../tests/yolo3_coco4/layers/c103.bin";
const char *c104_bin = "../tests/yolo3_coco4/layers/c104.bin";
const char *c105_bin = "../tests/yolo3_coco4/layers/c105.bin";
const char *g106_bin = "../tests/yolo3_coco4/layers/g106.bin";
const char *output_bins[3] = {
"../tests/yolo3_coco4/debug/layer82_out.bin",
"../tests/yolo3_coco4/debug/layer94_out.bin",
"../tests/yolo3_coco4/debug/layer106_out.bin"
};
int main() { int main() {
// Network layout // Network layout
tk::dnn::dataDim_t dim(1, 3, 416, 416, 1); tk::dnn::dataDim_t dim(1, 3, 416, 416, 1);
tk::dnn::Network net(dim); tk::dnn::Network net(dim);
tk::dnn::Conv2d c0 (&net, 32, 3, 3, 1, 1, 1, 1, c0_bin, true); // create yolo3 model
tk::dnn::Activation a0 (&net, tk::dnn::ACTIVATION_LEAKY); std::string bin_path = "../tests/yolo3_coco4";
tk::dnn::Conv2d c1 (&net, 64, 3, 3, 2, 2, 1, 1, c1_bin, true); int classes = 4;
tk::dnn::Activation a1 (&net, tk::dnn::ACTIVATION_LEAKY); tk::dnn::Yolo *yolo [3];
tk::dnn::Conv2d c2 (&net, 32, 1, 1, 1, 1, 0, 0, c2_bin, true); #include "models/Yolo3.h"
tk::dnn::Activation a2 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c3 (&net, 64, 3, 3, 1, 1, 1, 1, c3_bin, true);
tk::dnn::Activation a3 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s4 (&net, &a1);
tk::dnn::Conv2d c5 (&net, 128, 3, 3, 2, 2, 1, 1, c5_bin, true);
tk::dnn::Activation a5 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c6 (&net, 64, 1, 1, 1, 1, 0, 0, c6_bin, true);
tk::dnn::Activation a6 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c7 (&net, 128, 3, 3, 1, 1, 1, 1, c7_bin, true);
tk::dnn::Activation a7 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s8 (&net, &a5);
tk::dnn::Conv2d c9 (&net, 64, 1, 1, 1, 1, 0, 0, c9_bin, true);
tk::dnn::Activation a9 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c10 (&net, 128, 3, 3, 1, 1, 1, 1, c10_bin, true);
tk::dnn::Activation a10 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s11 (&net, &s8);
tk::dnn::Conv2d c12 (&net, 256, 3, 3, 2, 2, 1, 1, c12_bin, true);
tk::dnn::Activation a12 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c13 (&net, 128, 1, 1, 1, 1, 0, 0, c13_bin, true);
tk::dnn::Activation a13 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c14 (&net, 256, 3, 3, 1, 1, 1, 1, c14_bin, true);
tk::dnn::Activation a14 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s15 (&net, &a12);
tk::dnn::Conv2d c16 (&net, 128, 1, 1, 1, 1, 0, 0, c16_bin, true);
tk::dnn::Activation a16 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c17 (&net, 256, 3, 3, 1, 1, 1, 1, c17_bin, true);
tk::dnn::Activation a17 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s18 (&net, &s15);
tk::dnn::Conv2d c19 (&net, 128, 1, 1, 1, 1, 0, 0, c19_bin, true);
tk::dnn::Activation a19 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c20 (&net, 256, 3, 3, 1, 1, 1, 1, c20_bin, true);
tk::dnn::Activation a20 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s21 (&net, &s18);
tk::dnn::Conv2d c22 (&net, 128, 1, 1, 1, 1, 0, 0, c22_bin, true);
tk::dnn::Activation a22 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c23 (&net, 256, 3, 3, 1, 1, 1, 1, c23_bin, true);
tk::dnn::Activation a23 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s24 (&net, &s21);
tk::dnn::Conv2d c25 (&net, 128, 1, 1, 1, 1, 0, 0, c25_bin, true);
tk::dnn::Activation a25 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c26 (&net, 256, 3, 3, 1, 1, 1, 1, c26_bin, true);
tk::dnn::Activation a26 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s27 (&net, &s24);
tk::dnn::Conv2d c28 (&net, 128, 1, 1, 1, 1, 0, 0, c28_bin, true);
tk::dnn::Activation a28 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c29 (&net, 256, 3, 3, 1, 1, 1, 1, c29_bin, true);
tk::dnn::Activation a29 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s30 (&net, &s27);
tk::dnn::Conv2d c31 (&net, 128, 1, 1, 1, 1, 0, 0, c31_bin, true);
tk::dnn::Activation a31 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c32 (&net, 256, 3, 3, 1, 1, 1, 1, c32_bin, true);
tk::dnn::Activation a32 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s33 (&net, &s30);
tk::dnn::Conv2d c34 (&net, 128, 1, 1, 1, 1, 0, 0, c34_bin, true);
tk::dnn::Activation a34 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c35 (&net, 256, 3, 3, 1, 1, 1, 1, c35_bin, true);
tk::dnn::Activation a35 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s36 (&net, &s33);
tk::dnn::Conv2d c37 (&net, 512, 3, 3, 2, 2, 1, 1, c37_bin, true);
tk::dnn::Activation a37 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c38 (&net, 256, 1, 1, 1, 1, 0, 0, c38_bin, true);
tk::dnn::Activation a38 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c39 (&net, 512, 3, 3, 1, 1, 1, 1, c39_bin, true);
tk::dnn::Activation a39 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s40 (&net, &a37);
tk::dnn::Conv2d c41 (&net, 256, 1, 1, 1, 1, 0, 0, c41_bin, true);
tk::dnn::Activation a41 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c42 (&net, 512, 3, 3, 1, 1, 1, 1, c42_bin, true);
tk::dnn::Activation a42 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s43 (&net, &s40);
tk::dnn::Conv2d c44 (&net, 256, 1, 1, 1, 1, 0, 0, c44_bin, true);
tk::dnn::Activation a44 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c45 (&net, 512, 3, 3, 1, 1, 1, 1, c45_bin, true);
tk::dnn::Activation a45 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s46 (&net, &s43);
tk::dnn::Conv2d c47 (&net, 256, 1, 1, 1, 1, 0, 0, c47_bin, true);
tk::dnn::Activation a47 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c48 (&net, 512, 3, 3, 1, 1, 1, 1, c48_bin, true);
tk::dnn::Activation a48 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s49 (&net, &s46);
tk::dnn::Conv2d c50 (&net, 256, 1, 1, 1, 1, 0, 0, c50_bin, true);
tk::dnn::Activation a50 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c51 (&net, 512, 3, 3, 1, 1, 1, 1, c51_bin, true);
tk::dnn::Activation a51 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s52 (&net, &s49);
tk::dnn::Conv2d c53 (&net, 256, 1, 1, 1, 1, 0, 0, c53_bin, true);
tk::dnn::Activation a53 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c54 (&net, 512, 3, 3, 1, 1, 1, 1, c54_bin, true);
tk::dnn::Activation a54 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s55 (&net, &s52);
tk::dnn::Conv2d c56 (&net, 256, 1, 1, 1, 1, 0, 0, c56_bin, true);
tk::dnn::Activation a56 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c57 (&net, 512, 3, 3, 1, 1, 1, 1, c57_bin, true);
tk::dnn::Activation a57 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s58 (&net, &s55);
tk::dnn::Conv2d c59 (&net, 256, 1, 1, 1, 1, 0, 0, c59_bin, true);
tk::dnn::Activation a59 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c60 (&net, 512, 3, 3, 1, 1, 1, 1, c60_bin, true);
tk::dnn::Activation a60 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s61 (&net, &s58);
tk::dnn::Conv2d c62 (&net,1024, 3, 3, 2, 2, 1, 1, c62_bin, true);
tk::dnn::Activation a62 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c63 (&net, 512, 1, 1, 1, 1, 0, 0, c63_bin, true);
tk::dnn::Activation a63 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c64 (&net,1024, 3, 3, 1, 1, 1, 1, c64_bin, true);
tk::dnn::Activation a64 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s65 (&net, &a62);
tk::dnn::Conv2d c66 (&net, 512, 1, 1, 1, 1, 0, 0, c66_bin, true);
tk::dnn::Activation a66 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c67 (&net,1024, 3, 3, 1, 1, 1, 1, c67_bin, true);
tk::dnn::Activation a67 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s68 (&net, &s65);
tk::dnn::Conv2d c69 (&net, 512, 1, 1, 1, 1, 0, 0, c69_bin, true);
tk::dnn::Activation a69 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c70 (&net,1024, 3, 3, 1, 1, 1, 1, c70_bin, true);
tk::dnn::Activation a70 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s71 (&net, &s68);
tk::dnn::Conv2d c72 (&net, 512, 1, 1, 1, 1, 0, 0, c72_bin, true);
tk::dnn::Activation a72 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c73 (&net,1024, 3, 3, 1, 1, 1, 1, c73_bin, true);
tk::dnn::Activation a73 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Shortcut s74 (&net, &s71);
tk::dnn::Conv2d c75 (&net, 512, 1, 1, 1, 1, 0, 0, c75_bin, true);
tk::dnn::Activation a75 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c76 (&net,1024, 3, 3, 1, 1, 1, 1, c76_bin, true);
tk::dnn::Activation a76 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c77 (&net, 512, 1, 1, 1, 1, 0, 0, c77_bin, true);
tk::dnn::Activation a77 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c78 (&net,1024, 3, 3, 1, 1, 1, 1, c78_bin, true);
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, 27, 1, 1, 1, 1, 0, 0, c81_bin, false);
tk::dnn::Yolo yolo0 (&net, 4, 3, g82_bin);
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::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);
tk::dnn::Activation a88 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c89 (&net, 256, 1, 1, 1, 1, 0, 0, c89_bin, true);
tk::dnn::Activation a89 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c90 (&net, 512, 3, 3, 1, 1, 1, 1, c90_bin, true);
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, 27, 1, 1, 1, 1, 0, 0, c93_bin, false);
tk::dnn::Yolo yolo1 (&net, 4, 3, g94_bin);
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::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);
tk::dnn::Activation a100 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c101 (&net, 128, 1, 1, 1, 1, 0, 0, c101_bin, true);
tk::dnn::Activation a101 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c102 (&net, 256, 3, 3, 1, 1, 1, 1, c102_bin, true);
tk::dnn::Activation a102 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c103 (&net, 128, 1, 1, 1, 1, 0, 0, c103_bin, true);
tk::dnn::Activation a103 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c104 (&net, 256, 3, 3, 1, 1, 1, 1, c104_bin, true);
tk::dnn::Activation a104 (&net, tk::dnn::ACTIVATION_LEAKY);
tk::dnn::Conv2d c105 (&net, 27, 1, 1, 1, 1, 0, 0, c105_bin, false);
tk::dnn::Yolo yolo2 (&net, 4, 3, g106_bin);
// Load input // Load input
dnnType *data; dnnType *data;
@@ -304,9 +27,7 @@ int main() {
// the network have 3 outputs // the network have 3 outputs
tk::dnn::dataDim_t out_dim[3]; tk::dnn::dataDim_t out_dim[3];
out_dim[0] = yolo0.output_dim; for(int i=0; i<3; i++) out_dim[i] = yolo[i]->output_dim;
out_dim[1] = yolo1.output_dim;
out_dim[2] = yolo2.output_dim;
dnnType *cudnn_out[3], *rt_out[3]; dnnType *cudnn_out[3], *rt_out[3];
tk::dnn::dataDim_t dim1 = dim; //input dim tk::dnn::dataDim_t dim1 = dim; //input dim
@@ -317,18 +38,13 @@ int main() {
TIMER_STOP TIMER_STOP
dim1.print(); dim1.print();
} }
cudnn_out[0] = yolo0.dstData; for(int i=0; i<3; i++) cudnn_out[i] = yolo[i]->dstData;
cudnn_out[1] = yolo1.dstData;
cudnn_out[2] = yolo2.dstData;
printCenteredTitle(" compute detections ", '=', 30); printCenteredTitle(" compute detections ", '=', 30);
TIMER_START TIMER_START
int ndets = 0; int ndets = 0;
int classes = yolo0.classes;
tk::dnn::Yolo::detection *dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes); tk::dnn::Yolo::detection *dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes);
yolo0.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5); for(int i=0; i<3; i++) yolo[i]->computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
yolo1.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
yolo2.computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
tk::dnn::Yolo::mergeDetections(dets, ndets, classes); tk::dnn::Yolo::mergeDetections(dets, ndets, classes);
for(int j=0; j<ndets; j++) { for(int j=0; j<ndets; j++) {
@@ -356,9 +72,7 @@ int main() {
TIMER_STOP TIMER_STOP
dim2.print(); dim2.print();
} }
rt_out[0] = (dnnType*)netRT.buffersRT[1]; for(int i=0; i<3; i++) rt_out[i] = (dnnType*)netRT.buffersRT[i+1];
rt_out[1] = (dnnType*)netRT.buffersRT[2];
rt_out[2] = (dnnType*)netRT.buffersRT[3];
for(int i=0; i<3; i++) { for(int i=0; i<3; i++) {
printCenteredTitle((std::string(" YOLO ") + std::to_string(i) + " CHECK RESULTS ").c_str(), '=', 30); printCenteredTitle((std::string(" YOLO ") + std::to_string(i) + " CHECK RESULTS ").c_str(), '=', 30);
tests/yolo3_flir/yolo3_flir.cfg
+785
View File
@@ -0,0 +1,785 @@
[net]
# Testing
#batch=1
#subdivisions=1
# Training
batch=32
subdivisions=8
width=544
height=320
channels=1
momentum=0.9
decay=0.0005
angle=0
saturation = 1.5
exposure = 1.5
hue=.1
learning_rate=0.001
burn_in=1000
max_batches = 20000
policy=steps
steps=8000,9000
scales=.1,.1
[convolutional]
batch_normalize=1
filters=32
size=3
stride=1
pad=1
activation=leaky
# Downsample
[convolutional]
batch_normalize=1
filters=64
size=3
stride=2
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=32
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=64
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
# Downsample
[convolutional]
batch_normalize=1
filters=128
size=3
stride=2
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=64
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=128
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=64
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=128
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
# Downsample
[convolutional]
batch_normalize=1
filters=256
size=3
stride=2
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
# Downsample
[convolutional]
batch_normalize=1
filters=512
size=3
stride=2
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
# Downsample
[convolutional]
batch_normalize=1
filters=1024
size=3
stride=2
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=1024
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=1024
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=1024
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
filters=1024
size=3
stride=1
pad=1
activation=leaky
[shortcut]
from=-3
activation=linear
######################
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=1024
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=1024
activation=leaky
[convolutional]
batch_normalize=1
filters=512
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=1024
activation=leaky
[convolutional]
size=1
stride=1
pad=1
filters=24
activation=linear
[yolo]
mask = 6,7,8
anchors = 8.2087,8.5515, 18.4134,20.3391, 40.2194,29.2990, 31.6137,69.2240, 69.8497,48.3838, 108.8817,76.6316, 96.5753,145.5743, 165.9182,117.4493, 215.7497,198.4648
classes=3
num=9
jitter=.3
ignore_thresh = .5
truth_thresh = 1
random=0
[route]
layers = -4
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[upsample]
stride=2
[route]
layers = -1, 61
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=512
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=512
activation=leaky
[convolutional]
batch_normalize=1
filters=256
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=512
activation=leaky
[convolutional]
size=1
stride=1
pad=1
filters=24
activation=linear
[yolo]
mask = 3,4,5
anchors = 8.2087,8.5515, 18.4134,20.3391, 40.2194,29.2990, 31.6137,69.2240, 69.8497,48.3838, 108.8817,76.6316, 96.5753,145.5743, 165.9182,117.4493, 215.7497,198.4648
classes=3
num=9
jitter=.3
ignore_thresh = .5
truth_thresh = 1
random=0
[route]
layers = -4
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[upsample]
stride=2
[route]
layers = -1, 36
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=256
activation=leaky
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=256
activation=leaky
[convolutional]
batch_normalize=1
filters=128
size=1
stride=1
pad=1
activation=leaky
[convolutional]
batch_normalize=1
size=3
stride=1
pad=1
filters=256
activation=leaky
[convolutional]
size=1
stride=1
pad=1
filters=24
activation=linear
[yolo]
mask = 0,1,2
anchors = 8.2087,8.5515, 18.4134,20.3391, 40.2194,29.2990, 31.6137,69.2240, 69.8497,48.3838, 108.8817,76.6316, 96.5753,145.5743, 165.9182,117.4493, 215.7497,198.4648
classes=3
num=9
jitter=.3
ignore_thresh = .5
truth_thresh = 1
random=0
tests/yolo3_flir/yolo3_flir.cpp
+93
View File
@@ -0,0 +1,93 @@
#include<iostream>
#include<vector>
#include "tkdnn.h"
int main() {
// Network layout
tk::dnn::dataDim_t dim(1, 1, 320, 544, 1);
tk::dnn::Network net(dim);
// create yolo3 model
std::string bin_path = "../tests/yolo3_flir";
int classes = 3;
tk::dnn::Yolo *yolo [3];
#include "models/Yolo3.h"
// fill classes names
for(int i=0; i<3; i++) {
yolo[i]->classesNames = {"person", "bike", "car"};
}
// Load input
dnnType *data;
dnnType *input_h;
readBinaryFile(input_bin, dim.tot(), &input_h, &data);
//print network model
net.print();
//convert network to tensorRT
tk::dnn::NetworkRT netRT(&net, "yolo3_flir.rt");
// the network have 3 outputs
tk::dnn::dataDim_t out_dim[3];
for(int i=0; i<3; i++) out_dim[i] = yolo[i]->output_dim;
dnnType *cudnn_out[3], *rt_out[3];
tk::dnn::dataDim_t dim1 = dim; //input dim
printCenteredTitle(" CUDNN inference ", '=', 30); {
dim1.print();
TIMER_START
net.infer(dim1, data);
TIMER_STOP
dim1.print();
}
for(int i=0; i<3; i++) cudnn_out[i] = yolo[i]->dstData;
printCenteredTitle(" compute detections ", '=', 30);
TIMER_START
int ndets = 0;
tk::dnn::Yolo::detection *dets = tk::dnn::Yolo::allocateDetections(tk::dnn::Yolo::MAX_DETECTIONS, classes);
for(int i=0; i<3; i++) yolo[i]->computeDetections(dets, ndets, net.input_dim.w, net.input_dim.h, 0.5);
tk::dnn::Yolo::mergeDetections(dets, ndets, classes);
for(int j=0; j<ndets; j++) {
tk::dnn::Yolo::box b = dets[j].bbox;
int x0 = (b.x-b.w/2.);
int x1 = (b.x+b.w/2.);
int y0 = (b.y-b.h/2.);
int y1 = (b.y+b.h/2.);
int cl = 0;
for(int c = 0; c < classes; ++c){
float prob = dets[j].prob[c];
if(prob > 0)
cl = c;
}
std::cout<<cl<<": "<<x0<<" "<<y0<<" "<<x1<<" "<<y1<<"\n";
}
TIMER_STOP
tk::dnn::dataDim_t dim2 = dim;
printCenteredTitle(" TENSORRT inference ", '=', 30); {
dim2.print();
TIMER_START
netRT.infer(dim2, data);
TIMER_STOP
dim2.print();
}
for(int i=0; i<3; i++) rt_out[i] = (dnnType*)netRT.buffersRT[i+1];
for(int i=0; i<3; i++) {
printCenteredTitle((std::string(" YOLO ") + std::to_string(i) + " CHECK RESULTS ").c_str(), '=', 30);
dnnType *out, *out_h;
int odim = out_dim[i].tot();
readBinaryFile(output_bins[i], odim, &out_h, &out);
std::cout<<"CUDNN vs correct"; checkResult(odim, cudnn_out[i], out);
std::cout<<"TRT vs correct"; checkResult(odim, rt_out[i], out);
std::cout<<"CUDNN vs TRT "; checkResult(odim, cudnn_out[i], rt_out[i]);
}
return 0;
}
tkDNNConfig.cmake
-10
View File
@@ -1,10 +0,0 @@
find_package(CUDA REQUIRED)
find_package(OpenCV REQUIRED)
find_library(NVINFER NAMES nvinfer)
if(NVINFER STREQUAL "NVINFER-NOTFOUND")
set(NVINFER_INCLUDES "/usr/local/nvidia/tensorrt/include/")
link_directories(/usr/local/nvidia/tensorrt/targets/x86_64-linux-gnu/lib/
/usr/local/cuda/targets/x86_64-linux/lib/)
endif()
set(tkDNN_INCLUDE_DIRS ${CUDA_INCLUDE_DIRS} ${OPENCV_INCLUDE_DIRS} ${NVINFER_INCLUDES})
set(tkDNN_LIBRARIES tkDNN kernels ${CUDA_LIBRARIES} ${CUDA_CUBLAS_LIBRARIES} -lcudnn -lnvinfer ${OpenCV_LIBS})