Compare commits
200 Commits
| Author | SHA1 | Date | |
|---|---|---|---|
| 615b4c8a52 | |||
| fe37bcdf1e | |||
| 6911752e3e | |||
| 443691414a | |||
| 1a1c54f364 | |||
| 10b7160677 | |||
| 4746121d43 | |||
| 4fa5d2c231 | |||
| c1c2173e4d | |||
| 03d39d991c | |||
| a9c0db0bf6 | |||
| 8a4d1cac17 | |||
| 146e144249 | |||
| 2f57ba1222 | |||
| da4f246157 | |||
| 33844c1ab2 | |||
| 7233b065a8 | |||
| 2fa9f691ab | |||
| f3f5daf3db | |||
| c2d73623e5 | |||
| c32a0be257 | |||
| 57d7743f7e | |||
| 6bf9179acc | |||
| b218b18a02 | |||
| cfb457fdec | |||
| aa5927d8a1 | |||
| 1af2b792b8 | |||
| d5ae26dfef | |||
| 35787cc771 | |||
| 92f3d1c548 | |||
| bb7d382d96 | |||
| 4a7b290fdf | |||
| 5a1c7fb83d | |||
| bbc4dda635 | |||
| de8b02fe50 | |||
| ca62784f57 | |||
| ec02c7292f | |||
| 77f031c0f4 | |||
| a038e966d9 | |||
| 8c629ebe7b | |||
| b1a3620061 | |||
| 5bbb3f3480 | |||
| 88e0f9393a | |||
| 2d62d2524c | |||
| 7f667af48f | |||
| 2bcf9ab53b | |||
| ed83dfd99b | |||
| 041968f38a | |||
| f50aa4ad1a | |||
| 6656c3d0e8 | |||
| 5f444825ad | |||
| 12fd8d0109 | |||
| 32b6d51949 | |||
| 86da302163 | |||
| 4ebbb6af2b | |||
| ff5e376873 | |||
| a6d19d3698 | |||
| e38d8e82ca | |||
| caf4ddbce2 | |||
| 8627c5feeb | |||
| be31ae10d2 | |||
| f51a35ac5a | |||
| 428858eaae | |||
| 68ecd15125 | |||
| 3311196edb | |||
| 5c7301f7f4 | |||
| d2456b1d8a | |||
| eef1fd321f | |||
| 7d1d31ac45 | |||
| 1b9fe1ea61 | |||
| 753699104a | |||
| 978833fd6e | |||
| 17c5b7a818 | |||
| ca9d18c69e | |||
| 41ba8afa6d | |||
| ea1f0cc193 | |||
| 9b413b77ab | |||
| 9a4a65a3c3 | |||
| bd45b016bb | |||
| 28c012cade | |||
| 9b03bfcbd7 | |||
| 3c32d0c876 | |||
| 40c67e8536 | |||
| 38a1956404 | |||
| 21a698bb63 | |||
| a85367fa22 | |||
| 46c32edb94 | |||
| 3714155809 | |||
| 1c4aa3c5d7 | |||
| d631169821 | |||
| c22219ad16 | |||
| 61b6621d2c | |||
| 39df47574e | |||
| 1c8122f22d | |||
| 7505c28d2d | |||
| 39f80bbfb6 | |||
| 851c6a366c | |||
| de04ae1cab | |||
| 1aa4f0275d | |||
| c7941666ec | |||
| 87fe342ca2 | |||
| bdd8e0bc26 | |||
| 738fa94150 | |||
| 13063b904d | |||
| 0d682136de | |||
| 2c63bf05be | |||
| 0e97452460 | |||
| c8dea4668d | |||
| 88097a3774 | |||
| 2e8d0b1002 | |||
| 3bd725801d | |||
| 34be4cd00f | |||
| 3b60de00f8 | |||
| 53b429551d | |||
| 64626bf547 | |||
| 7a51b4382d | |||
| c13bda3863 | |||
| 2606820300 | |||
| a41b22e1f2 | |||
| c8f2e1b448 | |||
| 2ab47b5874 | |||
| 67cc566a0d | |||
| 217ff20058 | |||
| 991abdb410 | |||
| 7a46601306 | |||
| e91db28756 | |||
| ed02930464 | |||
| 5f25e0b5f6 | |||
| bc0ea65766 | |||
| dc55874f14 | |||
| 70373d638b | |||
| a9970f43fb | |||
| 6eb63160c8 | |||
| 6249956469 | |||
| 443179359d | |||
| a13bc2f007 | |||
| 4d30f0abd7 | |||
| 415bd47697 | |||
| 09679d7bb6 | |||
| 029ad71673 | |||
| 6331724953 | |||
| b7d240ea6d | |||
| 2cf8d8f6fc | |||
| a26ef98d2d | |||
| 747fddab3f | |||
| b2d6dcd207 | |||
| ab45c24efc | |||
| e449209d01 | |||
| e93ed59c30 | |||
| 168a1d8b27 | |||
| 6c2f6bcf2e | |||
| 030e14d782 | |||
| 00355cfcf4 | |||
| 0119b31455 | |||
| 37b050a9c8 | |||
| c41a0a09a6 | |||
| 5595b8037b | |||
| 5a52de17eb | |||
| 0aa9de4ce8 | |||
| c63ac6b590 | |||
| 2b4b9b8e49 | |||
| 66ad6bb1d6 | |||
| fc9fb4f153 | |||
| 6110fffbb5 | |||
| b3a369dc29 | |||
| 81e5f6a97b | |||
| 2d7563d27c | |||
| 3b2f062dd9 | |||
| 57c9a6ec99 | |||
| 04f96048b6 | |||
| 3124f86878 | |||
| 9a6058ac4a | |||
| aef39f6144 | |||
| 266330009c | |||
| b75fa637cb | |||
| 1c6888f312 | |||
| 3215d5aab0 | |||
| d7ce952465 | |||
| 0a9957ba18 | |||
| 7d570c0df4 | |||
| 34198a4e8d | |||
| b20a2e2902 | |||
| 0ff47ad6ba | |||
| 4e189755cf | |||
| 858b3501fa | |||
| 2ef76209a1 | |||
| 4526e2767a | |||
| e8355cee67 | |||
| 300b0af5dd | |||
| 714bd5f757 | |||
| bed0b57fad | |||
| ed5e5d58b5 | |||
| 1cfe70365f | |||
| b94931f9f7 | |||
| 8e4b3c6c17 | |||
| 483ffefc35 | |||
| e7a6f1fb6c | |||
| 0887199880 | |||
| e83baae1a7 | |||
| e4df86a07c |
@@ -1,5 +1,19 @@
|
||||
*~
|
||||
demo/demo/data/img_crop/
|
||||
demo/demo/data/img_disparity/
|
||||
demo/demo/data/map/
|
||||
demo/demo/data/masks_orient/
|
||||
demo/demo/data/pmat_new/
|
||||
demo/demo/data/masks_v2/
|
||||
build/
|
||||
.vscode/
|
||||
*.bin
|
||||
*.pyc
|
||||
*.pyc
|
||||
*.prototxt
|
||||
*.caffemodel
|
||||
*.h5
|
||||
*.tar.gz
|
||||
*.weights
|
||||
*.zip
|
||||
.idea/
|
||||
*.hdf5
|
||||
|
||||
@@ -0,0 +1,6 @@
|
||||
[submodule "tracker_CLASS"]
|
||||
path = tracker_CLASS
|
||||
url = https://github.com/mive93/tracker_CLASS.git
|
||||
[submodule "masa_protocol"]
|
||||
path = masa_protocol
|
||||
url = https://git.hipert.unimore.it/rcavicchioli/masa_protocol.git
|
||||
@@ -1,17 +1,155 @@
|
||||
cmake_minimum_required(VERSION 2.8)
|
||||
cmake_minimum_required(VERSION 3.5)
|
||||
|
||||
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)
|
||||
|
||||
find_package(CUDA QUIET REQUIRED)
|
||||
# project specific flags
|
||||
if(DEBUG)
|
||||
add_definitions(-DDEBUG)
|
||||
endif()
|
||||
|
||||
cuda_include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS})
|
||||
cuda_add_library(kernels SHARED src/kernels/activation_elu.cu)
|
||||
|
||||
include_directories(${CMAKE_CURRENT_SOURCE_DIR}/include ${CUDA_INCLUDE_DIRS})
|
||||
add_library(tkDNN SHARED src/Layer.cpp src/LayerWgs.cpp
|
||||
src/Dense.cpp src/Activation.cpp src/Conv2d.cpp src/Conv3d.cpp src/Flatten.cpp src/MulAdd.cpp src/Pooling.cpp
|
||||
src/Network.cpp src/utils.cpp)
|
||||
target_link_libraries(tkDNN kernels ${CUDA_LIBRARIES} ${CUDA_CUBLAS_LIBRARIES} ${CUDA_TOOLKIT_ROOT_DIR}/lib/libcudnn.so)
|
||||
#-------------------------------------------------------------------------------
|
||||
# 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)
|
||||
set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} -DOPENCV")
|
||||
|
||||
include_directories(/usr/include/gdal)
|
||||
|
||||
#-------------------------------------------------------------------------------
|
||||
# Build Libraries
|
||||
#-------------------------------------------------------------------------------
|
||||
file(GLOB tkdnn_SRC "src/*.cpp")
|
||||
set(tkdnn_LIBS kernels ${CUDA_LIBRARIES} ${CUDA_CUBLAS_LIBRARIES} ${CUDNN_LIBRARIES} ${OpenCV_LIBS})
|
||||
|
||||
file(GLOB class_SRC "src/class_src/*.cpp")
|
||||
|
||||
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 /usr/include/python2.7)
|
||||
|
||||
set(class_LIBS ${OpenCV_LIBS} -lgdal yaml-cpp python2.7)
|
||||
|
||||
add_library(tkDNN SHARED ${tkdnn_SRC})
|
||||
target_link_libraries(tkDNN ${tkdnn_LIBS})
|
||||
|
||||
add_library(CLASS SHARED ${class_SRC})
|
||||
target_link_libraries(CLASS ${class_LIBS})
|
||||
|
||||
|
||||
#static
|
||||
#add_library(tkDNN_static STATIC ${tkdnn_SRC})
|
||||
#target_link_libraries(tkDNN_static ${tkdnn_LIBS})
|
||||
|
||||
add_executable(test_simple tests/simple/test_simple.cpp)
|
||||
target_link_libraries(test_simple tkDNN)
|
||||
|
||||
add_executable(test_mnist tests/mnist/test_mnist.cpp)
|
||||
target_link_libraries(test_mnist tkDNN)
|
||||
|
||||
add_executable(test_mnistRT tests/mnist/test_mnistRT.cpp)
|
||||
target_link_libraries(test_mnistRT tkDNN)
|
||||
|
||||
## YOLO NETS
|
||||
add_executable(test_yolo tests/yolo/yolo.cpp)
|
||||
target_link_libraries(test_yolo tkDNN)
|
||||
|
||||
add_executable(test_yolo_voc tests/yolo_voc/yolo_voc.cpp)
|
||||
target_link_libraries(test_yolo_voc tkDNN)
|
||||
|
||||
add_executable(test_yolo_tiny tests/yolo_tiny/yolo_tiny.cpp)
|
||||
target_link_libraries(test_yolo_tiny tkDNN)
|
||||
|
||||
add_executable(test_yolo_relu tests/yolo_relu/yolo_relu.cpp)
|
||||
target_link_libraries(test_yolo_relu tkDNN)
|
||||
|
||||
|
||||
add_executable(test_yolo_224 tests/yolo_224/yolo_224.cpp)
|
||||
target_link_libraries(test_yolo_224 tkDNN)
|
||||
|
||||
add_executable(test_yolo_berkeley tests/yolo_berkeley/yolo_berkeley.cpp)
|
||||
target_link_libraries(test_yolo_berkeley tkDNN)
|
||||
|
||||
add_executable(test_yolo3_coco4 tests/yolo3_coco4/yolo3_coco4.cpp)
|
||||
target_link_libraries(test_yolo3_coco4 tkDNN)
|
||||
|
||||
add_executable(test_yolo3_berkeley tests/yolo3_berkeley/yolo3_berkeley.cpp)
|
||||
target_link_libraries(test_yolo3_berkeley tkDNN)
|
||||
|
||||
add_executable(test_yolo3_tetrapack tests/yolo3_tetrapack/yolo3_tetrapack.cpp)
|
||||
target_link_libraries(test_yolo3_tetrapack tkDNN)
|
||||
|
||||
add_executable(test_yolo3_tetrapack_resize tests/yolo3_tetrapack_resize/yolo3_tetrapack_resize.cpp)
|
||||
target_link_libraries(test_yolo3_tetrapack_resize tkDNN)
|
||||
|
||||
add_executable(test_yolo3_BCDS6 tests/yolo3_BCDS6/yolo3_BCDS6.cpp)
|
||||
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)
|
||||
################################################################################
|
||||
|
||||
|
||||
add_executable(test_rtinference tests/test_rtinference/rtinference.cpp)
|
||||
target_link_libraries(test_rtinference tkDNN)
|
||||
|
||||
add_executable(yolo3_demo demo/demo/demo.cpp
|
||||
tracker_CLASS/c++/src/ekf.cpp
|
||||
tracker_CLASS/c++/src/trackutils.cpp
|
||||
tracker_CLASS/c++/src/plot.cpp
|
||||
tracker_CLASS/c++/src/tracker.cpp )
|
||||
|
||||
|
||||
target_link_libraries(yolo3_demo tkDNN CLASS)
|
||||
|
||||
|
||||
|
||||
|
||||
#-------------------------------------------------------------------------------
|
||||
# Install
|
||||
#-------------------------------------------------------------------------------
|
||||
#if (CMAKE_INSTALL_PREFIX_INITIALIZED_TO_DEFAULT)
|
||||
# set (CMAKE_INSTALL_PREFIX "${CMAKE_BINARY_DIR}/install"
|
||||
# CACHE PATH "default install path" FORCE)
|
||||
#endif()
|
||||
message("install dir:" ${CMAKE_INSTALL_PREFIX})
|
||||
install(DIRECTORY include/ DESTINATION include/)
|
||||
install(TARGETS tkDNN kernels DESTINATION lib)
|
||||
install(DIRECTORY "${CMAKE_CURRENT_SOURCE_DIR}/cmake/" # source 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()
|
||||
|
||||
add_executable(tkDNNtest tests/test.cpp)
|
||||
target_link_libraries(tkDNNtest tkDNN)
|
||||
@@ -1,15 +1,18 @@
|
||||
# tkDNN
|
||||
tkDNN is a Deep Neural Network library built with cuDNN primitives specifically thought to work on NVIDIA TK1 board.<br>
|
||||
tkDNN is a Deep Neural Network library built with cuDNN primitives specifically thought to work on NVIDIA TK1(and all successive) board.<br>
|
||||
The main scope is to do high performance inference on already trained models.
|
||||
Currently supports the following layers:
|
||||
|
||||
* Dense, fully interconnected
|
||||
* Activation (RELU, ELU, SIGMOID, TANH)
|
||||
* Convolutional 2D
|
||||
* Convolutional 3D
|
||||
* Max and Average Pooling
|
||||
* Flatten
|
||||
* Data preprocessing
|
||||
this branch actually work on every NVIDIA GPU that support the dependencies:
|
||||
* CUDA 10.0
|
||||
* CUDNN 7.603
|
||||
* TENSORRT 6.01
|
||||
* OPENCV 4.1
|
||||
|
||||
## Dependencies
|
||||
|
||||
```
|
||||
sudo apt install libgdal-dev libeigen3-dev python-matplotlib libyaml-cpp-dev libcereal-dev python2.7-dev
|
||||
```
|
||||
|
||||
## Workflow
|
||||
The recommended workflow follow these step:
|
||||
@@ -24,62 +27,32 @@ Build with cmake
|
||||
mkdir build
|
||||
cd build
|
||||
cmake ..
|
||||
# use -DTEST_DATA=False to skip dataset download
|
||||
make
|
||||
```
|
||||
during the cmake configuration it will be dowloaded the weights needed for running
|
||||
the tests
|
||||
|
||||
## Test
|
||||
There is a ready to use example on *test* directory, to try it you must generate the weights with Keras
|
||||
```
|
||||
cd tests
|
||||
python test_model.py
|
||||
```
|
||||
And then execute the inference on build directory
|
||||
```
|
||||
cd build
|
||||
./tkDNNtest
|
||||
```
|
||||
this should output the same prediction as Keras.
|
||||
Assumiung you have correctly builded the library these are the test ready to exec:
|
||||
* test_simple: a simple convolutional and dense network (CUDNN only)
|
||||
* test_mnist: the famous mnist netwok (CUDNN and TENSORRT)
|
||||
* test_mnistRT: the mnist network hardcoded in using tensorRT apis (TENSORRT only)
|
||||
* test_yolo: YOLO detection network (CUDNN and TENSORRT)
|
||||
* test_yolo_tiny: smaller version of YOLO (CUDNN and TENSRRT)
|
||||
* test_yolo3_berkeley: our yolo3 version trained with BDD100K dateset
|
||||
|
||||
## Simple example
|
||||
Here is a example of the entire workflow on a simple model.
|
||||
Using the following Keras model save it to a file
|
||||
```python
|
||||
model = Sequential()
|
||||
model.add(Reshape((20, 1), input_shape=(20)))
|
||||
model.add(Dense(256))
|
||||
model.compile()
|
||||
|
||||
# save model
|
||||
model.save("path/to/model.h5")
|
||||
## yolo3 berkeley demo detection
|
||||
For the live detection you need to precompile the tensorRT file by luncing the desidered network test, this is the recommended process:
|
||||
```
|
||||
|
||||
After the model is created the weights can be exported for tkDNN inference
|
||||
export TKDNN_MODE=FP16 # set the half floating point optimization
|
||||
rm yolo3_berkeley.rt # be sure to delete(or move) old tensorRT files
|
||||
./test_yolo3_berkeley # run the yolo test (is slow)
|
||||
# with f16 inference the result will be a bit incorrect
|
||||
```
|
||||
python weights_exporter model.h5 dense --output=weights/path
|
||||
this will genereate a yolo3_berkeley.rt file that can be used for live detection:
|
||||
```
|
||||
the exporter take as arguments, in order:
|
||||
* input model
|
||||
* layer type ["dense", "conv2d", conv3d"]
|
||||
* { layer type ["dense", "conv2d", conv3d"] for each layer to export }
|
||||
* optional argument --output define path where export weights
|
||||
|
||||
Then we can create a c++ program to do inference on tk1
|
||||
```c++
|
||||
#include<tkdnn.h> //library include
|
||||
|
||||
//Network object
|
||||
tkDNN::Network net;
|
||||
//input dimension
|
||||
tkDNN::dataDim_t dim(1, 20, 1, 1, 1);
|
||||
//Dense layer
|
||||
tkDNN::Dense d0(&net, dim, 256, "weights/path", "bias/path");
|
||||
|
||||
//here load the input data to CUDA
|
||||
//value_type is an alias of "float"
|
||||
value_type *data_d = [...]
|
||||
|
||||
//do inference
|
||||
value_type *output_d = d0.infer(dim, data_d);
|
||||
//dim will be updated with the output dimension
|
||||
```
|
||||
The result is finally stored on output_d in device memory.
|
||||
./yolo3_demo # launch detection on a demo video
|
||||
./yolo3_demo yolo3_berkeley.rt /dev/video0 # launch detection on device 0
|
||||
```
|
||||

|
||||
|
||||
@@ -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)
|
||||
@@ -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)
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Mon 06 May 2019 11:32:13 PM CEST"
|
||||
image_width: 1920
|
||||
image_height: 1080
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 1.6158690952190570e+03, 0., 9.4702812371722337e+02, 0.,
|
||||
1.6123979985757153e+03, 5.1995630055718266e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -4.0971199964304100e-01, 1.8755404192050384e-01,
|
||||
-5.3059322427743867e-03, -1.0380603625304912e-03, 0. ]
|
||||
avg_reprojection_error: 3.4351035832972515e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:32:53 AM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 4.7264390181579711e+03, 0., 1.5059850098280642e+03, 0.,
|
||||
4.6793092340700096e+03, 6.7300681982359868e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -4.2669569210605879e-01, 6.6337608795749903e-01,
|
||||
-1.3881256269106437e-03, 5.2468063845700682e-03, 0. ]
|
||||
avg_reprojection_error: 3.1312290189919406e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:14:44 AM CEST"
|
||||
image_width: 1920
|
||||
image_height: 1080
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 1.6902498656747011e+03, 0., 9.7959318966703324e+02, 0.,
|
||||
1.7552884617253583e+03, 5.3327953707582492e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -5.4891909767312119e-01, 2.5555919841568631e-01,
|
||||
-4.3831358875660656e-03, -1.3934378903760349e-02, 0. ]
|
||||
avg_reprojection_error: 1.1758482932800183e+00
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Mon 06 May 2019 11:32:13 PM CEST"
|
||||
image_width: 1920
|
||||
image_height: 1080
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 1.6158690952190570e+03, 0., 9.4702812371722337e+02, 0.,
|
||||
1.6123979985757153e+03, 5.1995630055718266e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -4.0971199964304100e-01, 1.8755404192050384e-01,
|
||||
-5.3059322427743867e-03, -1.0380603625304912e-03, 0. ]
|
||||
avg_reprojection_error: 3.4351035832972515e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 09:56:50 AM CEST"
|
||||
image_width: 1920
|
||||
image_height: 1080
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 1.6229477302581809e+03, 0., 1.0277357980566628e+03, 0.,
|
||||
1.6485741394129034e+03, 5.5596919291027621e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.7853584845653426e-01, 7.8553352913896368e-02,
|
||||
-6.5552938633907229e-03, -1.6436824648695104e-02, 0. ]
|
||||
avg_reprojection_error: 8.4629096638637347e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Fri 03 May 2019 11:56:13 PM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 5.8796921906556563e+03, 0., 1.3036708932691290e+03, 0.,
|
||||
5.9435402023228071e+03, 8.1110067822514861e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -5.4688862790206871e-01, 5.1913397860290666e-01,
|
||||
-2.1076612628273591e-03, 1.6869796115416984e-02, 0. ]
|
||||
avg_reprojection_error: 6.7667474319420251e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Sat 04 May 2019 12:00:38 AM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 4.6903033136815602e+03, 0., 1.6303445000881884e+03, 0.,
|
||||
4.7582671272189546e+03, 4.3596515032334111e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.4366857232996317e-01, 2.2799325522263861e-01,
|
||||
2.0765840315530557e-02, -4.0088654509745098e-03, 0. ]
|
||||
avg_reprojection_error: 3.9811872397860709e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:20:53 AM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 2.5005410461483498e+03, 0., 1.5319405824251596e+03, 0.,
|
||||
2.5001544574623872e+03, 7.8267345299919543e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.7379752112038928e-01, 1.6246299444310250e-01,
|
||||
8.0371978716752837e-04, -9.6108499236087584e-04, 0. ]
|
||||
avg_reprojection_error: 3.6334262234685299e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:37:47 AM CEST"
|
||||
image_width: 960
|
||||
image_height: 720
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 5.0439587680799593e+02, 0., 4.8997081391816727e+02, 0.,
|
||||
5.0714582349015507e+02, 3.5481348085748095e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -2.7916140864065331e-01, 6.5465070220501562e-02,
|
||||
-1.9231901334709591e-03, -2.6191562264760264e-03, 0. ]
|
||||
avg_reprojection_error: 5.7283635087126605e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:41:12 AM CEST"
|
||||
image_width: 960
|
||||
image_height: 720
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 5.1663651913150818e+02, 0., 4.7267297458218127e+02, 0.,
|
||||
5.1291090124818436e+02, 3.8505850298928243e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -2.8051872523046845e-01, 6.0895269981008610e-02,
|
||||
-9.7920840355269542e-03, -4.9804820350633240e-04, 0. ]
|
||||
avg_reprojection_error: 5.4967308787122626e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:50:02 AM CEST"
|
||||
image_width: 960
|
||||
image_height: 720
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 4.9724079419911664e+02, 0., 4.9277930193807083e+02, 0.,
|
||||
4.9700744926387819e+02, 3.6581239154403062e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -2.7582961261093608e-01, 6.6908017283259263e-02,
|
||||
-2.1580546593114500e-03, -1.7921711595441153e-03, 0. ]
|
||||
avg_reprojection_error: 3.7129088933918375e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:53:27 AM CEST"
|
||||
image_width: 960
|
||||
image_height: 720
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 4.9372152821507876e+02, 0., 4.7585791077351445e+02, 0.,
|
||||
4.9644139996881893e+02, 3.5961856724726260e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -2.9023109325424973e-01, 8.3150964750672046e-02,
|
||||
-6.1378621304345154e-04, 8.4481910933416999e-04, 0. ]
|
||||
avg_reprojection_error: 3.4691001942524069e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Sat 04 May 2019 12:35:58 AM CEST"
|
||||
image_width: 3840
|
||||
image_height: 2160
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 2.1723071272381276e+03, 0., 1.9718118689531000e+03, 0.,
|
||||
2.2377541672328439e+03, 9.3157209524899565e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.8516162509048857e-01, 1.8961757063227327e-01,
|
||||
1.8297248985443184e-02, -8.9166274086698288e-03, 0. ]
|
||||
avg_reprojection_error: 9.9886914863900311e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Sun 05 May 2019 08:52:13 PM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 2.9841357325808735e+03, 0., 1.5379802472694901e+03, 0.,
|
||||
2.9784613885271938e+03, 8.9330228722164566e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -5.1096356919758967e-01, 1.4543132746407733e-01,
|
||||
-3.1254001577433334e-02, -1.4769334036191385e-02, 0. ]
|
||||
avg_reprojection_error: 9.3544537534095662e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Sun 05 May 2019 09:24:33 PM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 2.4796388675592771e+03, 0., 1.5358283835422017e+03, 0.,
|
||||
2.4440198814655632e+03, 8.9911455540136217e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.9478938399374452e-01, 1.6288159087710818e-01,
|
||||
-1.8565610712959927e-02, -7.0112574756757643e-03, 0. ]
|
||||
avg_reprojection_error: 4.5805459213906724e-01
|
||||
@@ -0,0 +1,22 @@
|
||||
%YAML:1.0
|
||||
---
|
||||
calibration_time: "Tue 07 May 2019 10:03:44 AM CEST"
|
||||
image_width: 3072
|
||||
image_height: 1728
|
||||
board_width: 8
|
||||
board_height: 6
|
||||
square_size: 2.4799999237060547e+01
|
||||
flags: 0
|
||||
camera_matrix: !!opencv-matrix
|
||||
rows: 3
|
||||
cols: 3
|
||||
dt: d
|
||||
data: [ 2.6027348174982544e+03, 0., 1.4808496083807213e+03, 0.,
|
||||
2.6008830910556521e+03, 6.7577068120137187e+02, 0., 0., 1. ]
|
||||
distortion_coefficients: !!opencv-matrix
|
||||
rows: 5
|
||||
cols: 1
|
||||
dt: d
|
||||
data: [ -3.4912899377320661e-01, 1.5704840296202566e-01,
|
||||
6.4926875404798358e-03, 5.7293259996249049e-03, 0. ]
|
||||
avg_reprojection_error: 4.0040122960491076e-01
|
||||
|
After Width: | Height: | Size: 5.3 MiB |
|
After Width: | Height: | Size: 3.0 MiB |
|
After Width: | Height: | Size: 28 KiB |
|
After Width: | Height: | Size: 91 KiB |
|
After Width: | Height: | Size: 86 KiB |
|
After Width: | Height: | Size: 36 KiB |
|
After Width: | Height: | Size: 40 KiB |
|
After Width: | Height: | Size: 38 KiB |
|
After Width: | Height: | Size: 37 KiB |
|
After Width: | Height: | Size: 45 KiB |
|
After Width: | Height: | Size: 39 KiB |
|
After Width: | Height: | Size: 82 KiB |
|
After Width: | Height: | Size: 97 KiB |
|
After Width: | Height: | Size: 89 KiB |
|
After Width: | Height: | Size: 86 KiB |
|
After Width: | Height: | Size: 103 KiB |
|
After Width: | Height: | Size: 101 KiB |
|
After Width: | Height: | Size: 15 KiB |
|
After Width: | Height: | Size: 18 KiB |
|
After Width: | Height: | Size: 16 KiB |
|
After Width: | Height: | Size: 17 KiB |
|
After Width: | Height: | Size: 16 KiB |
|
After Width: | Height: | Size: 23 KiB |
|
After Width: | Height: | Size: 16 KiB |
|
After Width: | Height: | Size: 17 KiB |
|
After Width: | Height: | Size: 128 KiB |
|
After Width: | Height: | Size: 142 KiB |
|
After Width: | Height: | Size: 82 KiB |
|
After Width: | Height: | Size: 86 KiB |
|
After Width: | Height: | Size: 82 KiB |
|
After Width: | Height: | Size: 61 KiB |
|
After Width: | Height: | Size: 86 KiB |
|
After Width: | Height: | Size: 83 KiB |
@@ -0,0 +1,3 @@
|
||||
-7.327951322519857 -89.11104695972499 9201.012307315486
|
||||
-3.2739347567332553 -42.14373356189513 3668.5149400959867
|
||||
-0.0009310035308911926 -0.010684335913469737 0.9999999999999999
|
||||
@@ -0,0 +1,3 @@
|
||||
-5.575118981764042 19.15613757077127 7116.209593619794
|
||||
-3.2016502214672586 11.087267568165238 4660.384266971425
|
||||
-0.0007553414127200226 0.002676057142671632 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-0.6209787370656474 -57.192772931734694 6553.189887121747
|
||||
-0.655304496315561 -28.195100762686465 3423.351597211246
|
||||
-0.0001275987755308679 -0.007884524572727637 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
21.441694474392808 72.45500438933801 4652.4523150318
|
||||
9.195810184174585 36.76655098601718 1801.6321216143722
|
||||
0.0027471614789484795 0.009610371445379061 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
23.89713091621312 60.523880117656084 7778.591483347838
|
||||
11.708054761838232 41.819183783017614 1126.9348661897782
|
||||
0.0029807026524197224 0.008783281479831609 0.9999999999999999
|
||||
@@ -0,0 +1,3 @@
|
||||
-2.986713098145805 -21.641128011999868 6870.903103366085
|
||||
-2.2937864581163416 -14.410613254705751 4594.877638262097
|
||||
-0.0004752252182089383 -0.003112785793525571 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
0.797119441300711 35.495698816749574 2225.384878217832
|
||||
0.3510443326339413 59.809425152982726 1419.4212456173686
|
||||
-1.8197506232485688e-05 0.01820976131616732 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
2.213864892663308 -21.831117643802187 4060.61306610222
|
||||
0.9075230321117119 -7.525265027425343 1268.217122301812
|
||||
0.0005431549026467866 -0.0054930502425022745 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-8.65061527352736 -23.247834477892475 4009.9545358742134
|
||||
-2.2093309523396223 -8.398154406654037 1529.4606465341913
|
||||
-0.001952580524058866 -0.006089308376531831 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-8.665173353556622 -19.056833294770822 3827.027102449301
|
||||
-2.728678410673174 -6.695956964089541 1523.1971944608492
|
||||
-0.0020218867007343213 -0.004877696027397352 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-0.5624066154780419 -26.36968028701744 3384.7920236393156
|
||||
-0.6590389521668137 -8.709883019993336 1162.7049573736972
|
||||
-0.00023421305979212285 -0.006824826866955653 0.9999999999999999
|
||||
@@ -0,0 +1,3 @@
|
||||
-0.9691844827251948 -16.63913290283819 3991.384826153086
|
||||
-0.47672518258283236 -11.365071723693541 1770.7201067235435
|
||||
-0.0003215600914247826 -0.005142032883105236 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-1.5172087045737013 5.226213011053854 2493.822558690447
|
||||
-1.4892268949636678 4.265359557115566 3382.185465939972
|
||||
-0.000492856899218649 0.0016917382715537009 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
3.671630456405992 102.56885593938215 3570.381264275829
|
||||
1.5985035927893978 46.57491080817888 -1491.4874345505793
|
||||
0.0006446393888135914 0.027894339637393125 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
0.44666021794890903 -32.932455985363575 3557.76647014547
|
||||
1.0444994361599629 -37.23760709667856 4404.4517853552425
|
||||
0.00029114742043039743 -0.010721230871024154 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
0.1982518156851575 -13.22038303683631 595.3838470607648
|
||||
-0.4806350794284377 -6.567788488139482 598.1732743332492
|
||||
-0.00016595895699412456 -0.007499178419813009 1.0
|
||||
@@ -0,0 +1,3 @@
|
||||
-0.4122368700442484 -10.479982981650977 812.1821012303
|
||||
-0.6536846365005352 -4.951476039607947 512.0831604767536
|
||||
-0.00042630219831388434 -0.006003223898658603 1.0
|
||||
@@ -0,0 +1,603 @@
|
||||
#include <time.h>
|
||||
|
||||
#include "utils.h"
|
||||
#include "Yolo3Detection.h"
|
||||
#include "message.h"
|
||||
#include "visualization.h"
|
||||
#include "configuration.h"
|
||||
|
||||
#include "tracker.h"
|
||||
#include "../masa_protocol/include/send.hpp"
|
||||
#include "../masa_protocol/include/serialize.hpp"
|
||||
|
||||
// #include <assert.h>
|
||||
// #include <unistd.h>
|
||||
// #include <mutex>
|
||||
// #include <ctime>
|
||||
// #include <pthread.h>
|
||||
// #include <signal.h>
|
||||
// #include <chrono>
|
||||
// #include <math.h>
|
||||
// #include <typeinfo>
|
||||
// #include <iostream>
|
||||
|
||||
#define MAX_DETECT_SIZE 100
|
||||
|
||||
bool gRun;
|
||||
std::chrono::steady_clock::time_point local_clock_start;
|
||||
std::mutex mutexgRun;
|
||||
std::string obj_class[10]{"person", "car", "truck", "bus", "motor", "bike", "rider", "traffic light", "traffic sign", "train"};
|
||||
//mutex for some opencv operations
|
||||
std::mutex mutex_cv;
|
||||
Show_t updates;
|
||||
bool SAVE_RESULT = false;
|
||||
|
||||
void sig_handler(int signo)
|
||||
{
|
||||
std::cout << "request gateway stop\n";
|
||||
mutexgRun.lock();
|
||||
gRun = false;
|
||||
mutexgRun.unlock();
|
||||
}
|
||||
|
||||
void *readVideoCapture(void *x_void_ptr)
|
||||
{
|
||||
std::cout << "readVideoCapture start...\n";
|
||||
|
||||
Frame_t *info_f = (Frame_t *)x_void_ptr;
|
||||
mutex_cv.lock();
|
||||
cv::VideoCapture cap(info_f->input, cv::CAP_FFMPEG);
|
||||
mutex_cv.unlock();
|
||||
cv::Mat frame_loc, frame0;
|
||||
int frame_nbr_loc = 0;
|
||||
// bool to_show = false;
|
||||
if (!cap.isOpened())
|
||||
{
|
||||
mutexgRun.lock();
|
||||
gRun = false;
|
||||
mutexgRun.unlock();
|
||||
}
|
||||
else
|
||||
std::cout << "camera started\n";
|
||||
|
||||
// cap.set(cv::CAP_PROP_BUFFERSIZE,3);
|
||||
// std::cout<<"buf size: "<<cap.get(CV_CAP_PROP_BUFFERSIZE)<<std::endl;
|
||||
auto start_t = std::chrono::steady_clock::now();
|
||||
auto step_t = std::chrono::steady_clock::now();
|
||||
auto end_t = std::chrono::steady_clock::now();
|
||||
auto current_timestamp = std::chrono::steady_clock::now();
|
||||
|
||||
// compute fps and find camera's clock
|
||||
double shift, mean_time = 0;
|
||||
std::cout << "Frames per second using video.get(cv::CAP_PROP_FPS) : " << cap.get(cv::CAP_PROP_FPS) << std::endl;
|
||||
std::cout << "readVideoCapture computes frame rate...\n";
|
||||
// //compute frame rate
|
||||
int i = 0;
|
||||
int num_f = 120;
|
||||
// the first 20 frames are null
|
||||
while (i < 21)
|
||||
{
|
||||
cap >> frame_loc;
|
||||
i++;
|
||||
}
|
||||
|
||||
i = 0;
|
||||
start_t = std::chrono::steady_clock::now();
|
||||
while (i < num_f)
|
||||
{
|
||||
step_t = std::chrono::steady_clock::now();
|
||||
cap >> frame_loc;
|
||||
mean_time = mean_time + std::chrono::duration_cast<std::chrono::milliseconds>(std::chrono::steady_clock::now() - step_t).count();
|
||||
std::cout << " step " << i << " : " << std::chrono::duration_cast<std::chrono::milliseconds>(std::chrono::steady_clock::now() - step_t).count() << " ms" << std::endl;
|
||||
i++;
|
||||
}
|
||||
end_t = std::chrono::steady_clock::now();
|
||||
|
||||
std::cout << "Capturing " << num_f << " frames" << std::endl;
|
||||
std::cout << " Time taken : " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count() << " ms" << std::endl;
|
||||
|
||||
/*
|
||||
mean_time indicates the milliseconds from a frame and the next. (frame rate)
|
||||
local_clock_sync is the camera clock.
|
||||
shift is the difference from local camera clock and local process clock.
|
||||
a frame is allowed if its local timestamp minus its local clock is less then a tollerance,
|
||||
otherwise it will be considered old.
|
||||
*/
|
||||
auto local_clock_sync = std::chrono::steady_clock::now();
|
||||
mean_time = mean_time / num_f;
|
||||
shift = ((double)std::chrono::duration_cast<std::chrono::milliseconds>(local_clock_sync - local_clock_start).count()) / mean_time;
|
||||
shift = (shift - (int)shift) * mean_time;
|
||||
std::cout << ".-------------------------------\n";
|
||||
std::cout << " mean time: " << mean_time << std::endl;
|
||||
std::cout << " shift: " << shift << std::endl;
|
||||
std::cout << " TIMEDIFFERENCE: " << std::chrono::duration_cast<std::chrono::milliseconds>(local_clock_sync - local_clock_start).count() << std::endl;
|
||||
std::cout << "\n\n\n\n";
|
||||
std::cout << "readVideoCapture start to capture...\n";
|
||||
while (gRun)
|
||||
{
|
||||
// mutex_cv.lock();
|
||||
cap >> frame_loc;
|
||||
// mutex_cv.unlock();
|
||||
current_timestamp = std::chrono::steady_clock::now();
|
||||
shift = std::chrono::duration_cast<std::chrono::milliseconds>(current_timestamp - local_clock_sync).count();
|
||||
std::cout << " RELATIVE TIMESTAMP FRAME : " << shift << " ms" << std::endl;
|
||||
shift = shift / mean_time;
|
||||
shift = (shift - (int)shift) * mean_time;
|
||||
shift = (shift - mean_time / 2 >= 0) ? -(mean_time - shift) : shift;
|
||||
std::cout << "DELAY frame_" << frame_nbr_loc << " : " << shift << " ms" << std::endl;
|
||||
// TODO: here introduce a tollerance to discard old frame
|
||||
|
||||
// std::cout<< "CV_CAP_PROP_POS_MSEC: "<< cap.get( cv::CAP_PROP_POS_MSEC) <<std::endl;
|
||||
// std::cout<< "CV_CAP_PROP_POS_FRAMES: "<< cap.get( cv::CAP_PROP_POS_FRAMES) <<std::endl; // <-- the v4l2 'sequence' field
|
||||
// std::cout<< "cv::CAP_PROP_FPS: "<< cap.get( cv::CAP_PROP_FPS)<<std::endl;
|
||||
// std::cout << "Format: " << cap.get(CV_CAP_PROP_FORMAT) << "\n";
|
||||
// CAP_PROP_POS_MSEC Current position of the video file in milliseconds or video capture timestamp.
|
||||
std::cout << "id: " << cap.get(cv::CAP_PROP_POS_MSEC) << std::endl;
|
||||
// CAP_PROP_FRAME_COUNT Number of frames in the video file.
|
||||
std::cout << "id: " << cap.get(cv::CAP_PROP_FRAME_COUNT) << std::endl;
|
||||
|
||||
if (!frame_loc.data)
|
||||
{
|
||||
usleep(1000000);
|
||||
mutex_cv.lock();
|
||||
cap.open(info_f->input);
|
||||
printf("cap reinitialize\n");
|
||||
mutex_cv.unlock();
|
||||
continue;
|
||||
}
|
||||
|
||||
end_t = std::chrono::steady_clock::now();
|
||||
std::cout << " VC-TIME 1 : " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count() << " ms" << std::endl;
|
||||
start_t = end_t;
|
||||
|
||||
info_f->sem_vc.lock();
|
||||
info_f->frame = frame_loc.clone();
|
||||
info_f->frame_nbr = frame_nbr_loc;
|
||||
info_f->sem_vc.unlock();
|
||||
// usleep(50000);
|
||||
end_t = std::chrono::steady_clock::now();
|
||||
std::cout << " VC-TIME 2 : " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - start_t).count() << " ms" << std::endl;
|
||||
start_t = end_t;
|
||||
frame_nbr_loc++;
|
||||
}
|
||||
return (void *)0;
|
||||
}
|
||||
|
||||
void *computationTask(void *x_void_ptr)
|
||||
{
|
||||
Camera_t *camera = (Camera_t *)x_void_ptr;
|
||||
pthread_t visual, originalshow, detectionshow, topviewshow, disparityshow;
|
||||
pthread_t videocap;
|
||||
tk::dnn::Yolo3Detection yolo = *(camera->yolo);
|
||||
//create video capture thread
|
||||
Frame_t info_f;
|
||||
info_f.input = camera->input;
|
||||
if (pthread_create(&videocap, NULL, readVideoCapture, (void *)&info_f))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
|
||||
bool to_show = camera->to_show;
|
||||
double adfGeoTransform[6];
|
||||
for (int i = 0; i < 6; i++)
|
||||
adfGeoTransform[i] = camera->adfGeoTransform[i];
|
||||
|
||||
ModFrame_t info_show;
|
||||
if (to_show)
|
||||
{
|
||||
// initialize updates struct
|
||||
updates.update_o = false;
|
||||
updates.update_de = false;
|
||||
updates.update_t = false;
|
||||
updates.update_di = false;
|
||||
if (pthread_create(&visual, NULL, show_updates, (void *)NULL))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
if (pthread_create(&originalshow, NULL, originalFrame, (void *)&info_f))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
if (pthread_create(&disparityshow, NULL, disparityFrame, (void *)&info_f))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
info_show.H = cv::Mat(cv::Size(3, 3), CV_64FC1);
|
||||
if (pthread_create(&detectionshow, NULL, detectionFrame, (void *)&info_show))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
if (pthread_create(&topviewshow, NULL, topviewFrame, (void *)&info_show))
|
||||
{
|
||||
fprintf(stderr, "Error creating thread\n");
|
||||
return (void *)1;
|
||||
};
|
||||
}
|
||||
char *pmatrix = camera->pmatrix;
|
||||
/*projection matrix from camera to map*/
|
||||
cv::Mat H(cv::Size(3, 3), CV_64FC1);
|
||||
read_projection_matrix(H, pmatrix);
|
||||
assert(cv::countNonZero(H) > 0);
|
||||
// std::cout<<H<<std::endl;
|
||||
// return (void*)0;
|
||||
/*Camera calibration*/
|
||||
cv::Mat cameraMat, distCoeff;
|
||||
readCameraCalibrationYaml(camera->cameraCalib, cameraMat, distCoeff);
|
||||
std::cout << cameraMat << std::endl;
|
||||
std::cout << distCoeff << std::endl;
|
||||
|
||||
/*GPS information*/
|
||||
std::vector<ObjCoords> coords;
|
||||
|
||||
/*socket*/
|
||||
Communicator Comm(SOCK_DGRAM);
|
||||
Comm.open_client_socket((char *)"127.0.0.1", 8888);
|
||||
|
||||
Message *m = new Message;
|
||||
m->cam_idx = camera->CAM_IDX;
|
||||
m->lights.clear();
|
||||
/*Conversion for tracker, from gps to meters and viceversa*/
|
||||
// mutex_cv.lock();
|
||||
geodetic_converter::GeodeticConverter gc;
|
||||
gc.initialiseReference(44.655540, 10.934315, 0);
|
||||
// mutex_cv.unlock();
|
||||
double east, north, up;
|
||||
// double lat, lon, alt;
|
||||
/*Mask info*/
|
||||
cv::Mat mask = cv::imread(camera->maskfile, cv::IMREAD_GRAYSCALE);
|
||||
cv::Mat maskOrient = cv::imread(camera->maskFileOrient);
|
||||
// cv::Mat maskOrient = cv::imread(camera->maskFileOrient, 0);
|
||||
|
||||
/*for(int i=0; i< mask.cols; i++)
|
||||
{
|
||||
for(int j=0; j< mask.rows; j++)
|
||||
{
|
||||
std::cout<<maskOrient.at<cv::Vec3b>(i,j) <<std::endl;
|
||||
}
|
||||
}
|
||||
|
||||
|
||||
return 0;*/
|
||||
/*tracker infos*/
|
||||
std::vector<Tracker> trackers;
|
||||
std::vector<Data> cur_frame;
|
||||
int initial_age = -5;
|
||||
int age_threshold = -8;
|
||||
int n_states = 5;
|
||||
float dt = 0.03;
|
||||
|
||||
int frame_nbr = 0;
|
||||
|
||||
//save video
|
||||
/*cv::VideoWriter outputVideo;
|
||||
cv::Size S = cv::Size((int)cap.get(cv::CAP_PROP_FRAME_WIDTH), //Acquire input size
|
||||
(int)cap.get(cv::CAP_PROP_FRAME_HEIGHT));
|
||||
outputVideo.open("test.avi", static_cast<int>(cap.get(cv::CAP_PROP_FOURCC)), cap.get(cv::CAP_PROP_FPS), S, true);*/
|
||||
|
||||
cv::Mat map1, map2;
|
||||
auto start_t = std::chrono::steady_clock::now();
|
||||
auto step_t = std::chrono::steady_clock::now();
|
||||
auto end_t = std::chrono::steady_clock::now();
|
||||
// auto step_t_segmentation = std::chrono::steady_clock::now();
|
||||
// auto end_t_segmentation = std::chrono::steady_clock::now();
|
||||
|
||||
//TODO: move in a thread
|
||||
// // information for the disparity map
|
||||
// std::vector <cv::Rect> pre_rois;
|
||||
// cv::Mat pre_frame;
|
||||
cv::Mat orig_frame;
|
||||
// cv::Mat canny, pre_canny, canny_RGB, pre_canny_RGB;
|
||||
// cv::Mat canny_img;
|
||||
|
||||
// box variable
|
||||
tk::dnn::box b;
|
||||
int x0, h, y1; //w, x1, y0;
|
||||
int objClass;
|
||||
std::string det_class;
|
||||
;
|
||||
// float prob;
|
||||
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_crop;
|
||||
cv::Mat dnn_input;
|
||||
bool first_iteration = true;
|
||||
|
||||
while (gRun)
|
||||
{
|
||||
TIMER_START
|
||||
start_t = std::chrono::steady_clock::now();
|
||||
step_t = start_t;
|
||||
|
||||
info_f.sem_vc.lock();
|
||||
frame = info_f.frame.clone();
|
||||
if (info_f.frame_nbr - frame_nbr > 1)
|
||||
std::cout << "more than one - f_n (diff " << info_f.frame_nbr - frame_nbr << ")\n";
|
||||
frame_nbr = info_f.frame_nbr;
|
||||
info_f.sem_vc.unlock();
|
||||
std::cout << "f_n: " << frame_nbr << std::endl;
|
||||
// if (!frame.data)
|
||||
if (frame_nbr == 0)
|
||||
{
|
||||
usleep(1000000);
|
||||
printf("no frame received\n");
|
||||
continue;
|
||||
}
|
||||
orig_frame = frame.clone();
|
||||
// mutex_cv.lock();
|
||||
if (first_iteration)
|
||||
cv::initUndistortRectifyMap(cameraMat, distCoeff, cv::Mat(), cameraMat, frame.size(), CV_16SC2, map1, map2);
|
||||
cv::Mat temp = frame.clone();
|
||||
cv::remap(temp, frame, map1, map2, 1);
|
||||
//undistort(temp, frame, cameraMat, distCoeff);
|
||||
// mutex_cv.unlock();
|
||||
|
||||
// this will be resized to the net format
|
||||
dnn_input = frame.clone();
|
||||
// TODO: async infer
|
||||
yolo.update(dnn_input);
|
||||
int num_detected = yolo.detected.size();
|
||||
if (num_detected > MAX_DETECT_SIZE)
|
||||
num_detected = MAX_DETECT_SIZE;
|
||||
|
||||
coords.clear();
|
||||
end_t = std::chrono::steady_clock::now();
|
||||
std::cout << " TIME 1 : " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - step_t).count() << " ms" << std::endl;
|
||||
step_t = end_t;
|
||||
// draw dets
|
||||
std::cout << "camera: " << camera->CAM_IDX << " - num detected: " << num_detected << std::endl;
|
||||
|
||||
//TODO: move in a thread
|
||||
// //preprocessing frame
|
||||
// step_t_segmentation = std::chrono::steady_clock::now();
|
||||
// // src_gray
|
||||
// canny_img = img_laplacian(orig_frame,0);
|
||||
// cv::Canny(canny_img, canny, 100, 100*2 );
|
||||
// // sprintf(buf_frame_crop_name,"../demo/demo/data/img_disparity/%d_%d_canny.jpg",frame_nbr, 999);
|
||||
// // cv::imwrite(buf_frame_crop_name, canny);
|
||||
// end_t_segmentation = std::chrono::steady_clock::now();
|
||||
// std::cout << " - TIME END pre canny : "<<std::chrono::duration_cast<std::chrono::milliseconds>(end_t_segmentation - step_t_segmentation).count() << " ms"<<std::endl;
|
||||
// step_t_segmentation = end_t_segmentation;
|
||||
// // std::cout<<"o: "<<orig_frame.cols<<" - "<<orig_frame.rows<<std::endl;
|
||||
// // std::cout<<"canny: "<<canny.cols<<" - "<<canny.rows<<std::endl;
|
||||
// // std::cout<<"pre: "<<pre_canny.cols<<" - "<<pre_canny.rows<<std::endl;
|
||||
// if(!first_iteration)
|
||||
// {
|
||||
// // backtorgb = cv::cvtColor(pre_canny,cv::COLOR_GRAY2RGB)
|
||||
// cv::cvtColor(pre_canny, pre_canny_RGB, cv::COLOR_GRAY2RGB);
|
||||
// cv::cvtColor(canny, canny_RGB, cv::COLOR_GRAY2RGB);
|
||||
// disparity_frame = frame_disparity(pre_canny_RGB, canny_RGB, frame_nbr, 999, 0);
|
||||
// std::cout<<"size: "<<disparity_frame.rows<<" - "<<disparity_frame.cols<<std::endl;
|
||||
// if (disparity_frame.rows == 0 || disparity_frame.cols == 0)
|
||||
// return -1;
|
||||
// if (disparity_frame.empty())
|
||||
// { // only fools don't check...
|
||||
// std::cout << "image not loaded !" << std::endl;
|
||||
// return -1;
|
||||
// }
|
||||
// end_t_segmentation = std::chrono::steady_clock::now();
|
||||
// std::cout << " TIME canny : frame_disparity : "<<std::chrono::duration_cast<std::chrono::milliseconds>(end_t_segmentation - step_t_segmentation).count() << " ms"<<std::endl;
|
||||
// step_t_segmentation = end_t_segmentation;
|
||||
|
||||
// // //--------------------------------
|
||||
// // //frame box disparity on the original image
|
||||
// // step_t_segmentation = std::chrono::steady_clock::now();
|
||||
// // frame_box_disparity(pre_frame, frame, pre_rois, frame_nbr);
|
||||
// // // reset pre_rois for the new roi of the current frame
|
||||
// // // pre_rois.erase(pre_rois.begin(), pre_rois.end());
|
||||
// // end_t_segmentation = std::chrono::steady_clock::now();
|
||||
// // std::cout << " TIME Frame disparity : "<<std::chrono::duration_cast<std::chrono::milliseconds>(end_t_segmentation - step_t_segmentation).count() << " ms"<<std::endl;
|
||||
// // step_t_segmentation = end_t_segmentation;
|
||||
|
||||
// // //frame box disparity on the preprocessed image
|
||||
// // cv::cvtColor(pre_canny, pre_canny_RGB, cv::COLOR_GRAY2RGB);
|
||||
// // cv::cvtColor(canny, canny_RGB, cv::COLOR_GRAY2RGB);
|
||||
// // frame_box_disparity(pre_canny_RGB, canny_RGB, pre_rois, frame_nbr);
|
||||
// // // reset pre_rois for the new roi of the current frame
|
||||
// // pre_rois.erase(pre_rois.begin(), pre_rois.end());
|
||||
// // end_t_segmentation = std::chrono::steady_clock::now();
|
||||
// // std::cout << " TIME Canny Frame disparity : "<<std::chrono::duration_cast<std::chrono::milliseconds>(end_t_segmentation - step_t_segmentation).count() << " ms"<<std::endl;
|
||||
// // step_t_segmentation = end_t_segmentation;
|
||||
// // //---------------------------------
|
||||
// }
|
||||
|
||||
// compute some metrics on the whole frame
|
||||
// segmentation(pre_frame, frame, frame_nbr, 0, 0);
|
||||
|
||||
for (int i = 0; i < num_detected; i++)
|
||||
{
|
||||
b = yolo.detected[i];
|
||||
x0 = b.x;
|
||||
// w = b.w;
|
||||
// x1 = b.x + w;
|
||||
// y0 = b.y;
|
||||
h = b.h;
|
||||
y1 = b.y + h;
|
||||
objClass = b.cl;
|
||||
det_class = obj_class[b.cl];
|
||||
// prob = b.prob;
|
||||
|
||||
intensity = mask.at<uchar>(cv::Point(int(x0 + b.w / 2), y1));
|
||||
|
||||
if (intensity[0])
|
||||
{
|
||||
|
||||
if (objClass < 6)
|
||||
{
|
||||
|
||||
// find the rectangular on the frame (sub-figure)
|
||||
// roi.x = (x0 > 0)? x0 : 0;
|
||||
// roi.y = (y0 > 0)? y0 : 0;
|
||||
// // std::cout<<"x "<<roi.x<<" - y "<<roi.y<<std::endl;
|
||||
// roi.width = (roi.x+w >= frame.cols)? frame.cols-1-roi.x : w;
|
||||
// roi.height = (roi.y+h >= frame.rows)? frame.rows-1-roi.y : h;
|
||||
// std::cout<<"w "<<roi.width<<" - h "<<roi.height<<std::endl;
|
||||
// std::cout<<"wf "<<frame.cols<<" - hf "<<frame.rows<<std::endl;
|
||||
// std::cout<<"---"<<std::endl;
|
||||
// std::cout<<"x "<<roi.x<<" to "<<roi.width+roi.x<<" wf "<<frame.cols<<std::endl;
|
||||
// std::cout<<"y "<<roi.y<<" to "<<roi.height+roi.y<<" hf "<<frame.rows<<std::endl;
|
||||
//update pre_roi for the next frame
|
||||
// pre_rois.push_back(roi);
|
||||
|
||||
// segmentation(frame(roi), frame(roi), frame_nbr, i, 1);
|
||||
|
||||
/////
|
||||
convert_coords(coords, x0 + b.w / 2, y1, objClass, H, adfGeoTransform);
|
||||
|
||||
// //std::cout<<objClass<<" ("<<prob<<"): "<<x0<<" "<<y0<<" "<<x1<<" "<<y1<<"\n";
|
||||
// cv::rectangle(frame, cv::Point(x0, y0), cv::Point(x1, y1), camera->yolo.colors[objClass], 2);
|
||||
// // draw label
|
||||
// int baseline = 0;
|
||||
// float fontScale = 0.5;
|
||||
// int thickness = 2;
|
||||
// cv::Size textSize = getTextSize(det_class, cv::FONT_HERSHEY_SIMPLEX, fontScale, thickness, &baseline);
|
||||
// cv::rectangle(frame, cv::Point(x0, y0), cv::Point((x0 + textSize.width - 2), (y0 - textSize.height - 2)), camera->yolo.colors[b.cl], -1);
|
||||
// cv::putText(frame, det_class, cv::Point(x0, (y0 - (baseline / 2))), cv::FONT_HERSHEY_SIMPLEX, fontScale, cv::Scalar(255, 255, 255), thickness);
|
||||
}
|
||||
}
|
||||
}
|
||||
|
||||
end_t = std::chrono::steady_clock::now();
|
||||
std::cout << " TIME 2 : " << std::chrono::duration_cast<std::chrono::milliseconds>(end_t - step_t).count() << " ms" << std::endl;
|
||||
step_t = end_t;
|
||||
//convert from latitude and longitude to meters for ekf
|
||||
cur_frame.clear();
|
||||
for (size_t i = 0; i < coords.size(); i++)
|
||||
{
|
||||
gc.geodetic2Enu(coords[i].lat_, coords[i].long_, 0, &east, &north, &up);
|
||||
cur_frame.push_back(Data(east, north, frame_nbr, coords[i].class_));
|
||||
}
|
||||
if (first_iteration)
|
||||
{
|
||||
// if there aren't detections and it is the first iteration, we can't initialize the tracker, so continue
|
||||
if (cur_frame.empty())
|
||||
continue;
|
||||
for (auto f : cur_frame)
|
||||
trackers.push_back(Tracker(f, initial_age, dt, n_states));
|
||||
}
|
||||
else
|
||||
{
|
||||
Track(cur_frame, dt, n_states, initial_age, age_threshold, trackers);
|
||||
}
|
||||
std::cout << "There are " << trackers.size() << " trackers" << std::endl;
|
||||
//prepare message with tracker info
|
||||
if (trackers.size() != 0)
|
||||
{
|
||||
// mutex_cv.lock();
|
||||
addRoadUserfromTracker(trackers, m, gc, maskOrient, adfGeoTransform, H);
|
||||
// mutex_cv.unlock();
|
||||
//prepare the message with detection info
|
||||
//prepare_message(m, coords, CAM_IDX);
|
||||
//send message
|
||||
if (!m->objects.empty())
|
||||
Comm.send_message(m);
|
||||
}
|
||||
|
||||
if (to_show)
|
||||
{
|
||||
//populate the ModFrame_t
|
||||
info_show.sem.lock();
|
||||
info_show.original_frame = frame.clone();
|
||||
// std::vector<Tracker> trackers;
|
||||
info_show.trackers = trackers;
|
||||
// geodetic_converter::GeodeticConverter gc;
|
||||
info_show.gc = gc;
|
||||
for (int i = 0; i < 6; i++)
|
||||
info_show.adfGeoTransform[i] = adfGeoTransform[i];
|
||||
// cv::Mat H;
|
||||
info_show.H = H.clone();
|
||||
info_show.yolo = yolo;
|
||||
// std::copy(camera->yolo.begin(), camera->yolo.end(), info_show.yolo.begin());
|
||||
info_show.mask = mask.clone();
|
||||
info_show.sem.unlock();
|
||||
}
|
||||
|
||||
// update pre_frame for the disparity map
|
||||
// pre_frame = orig_frame.clone();
|
||||
// pre_canny = canny.clone();
|
||||
if (first_iteration)
|
||||
first_iteration = false;
|
||||
|
||||
frame_nbr++;
|
||||
std::cout << camera->CAM_IDX << " camera thread: ";
|
||||
TIMER_STOP
|
||||
}
|
||||
return (void *)0;
|
||||
}
|
||||
|
||||
int main(int argc, char *argv[])
|
||||
{
|
||||
|
||||
std::cout << "detection\n";
|
||||
signal(SIGINT, sig_handler);
|
||||
srand(time(NULL));
|
||||
|
||||
Parameters_t par;
|
||||
|
||||
if(!read_parameters(argc, argv, &par))
|
||||
return -1;
|
||||
|
||||
tk::dnn::Yolo3Detection yolo[par.n_cameras];
|
||||
for(int i=0; i<par.n_cameras; i++)
|
||||
{
|
||||
yolo[i].init(par.net);
|
||||
yolo[i].thresh = 0.25;
|
||||
|
||||
// if(SAVE_RESULT)
|
||||
// resultVideo << frame;
|
||||
}
|
||||
// tk::dnn::Yolo3Detection yolo;
|
||||
// yolo.init(net);
|
||||
// yolo.thresh = 0.25;
|
||||
|
||||
gRun = true;
|
||||
// start the local clock. It is used to check the incoming frames (by different cameras)
|
||||
local_clock_start = std::chrono::steady_clock::now();
|
||||
|
||||
/*GPS information*/
|
||||
double *adfGeoTransform = (double *)malloc(6 * sizeof(double));
|
||||
readTiff(par.tiffile, adfGeoTransform);
|
||||
// Camera_t cameras[par.n_cameras];
|
||||
for(int i=0; i<par.n_cameras; i++)
|
||||
{
|
||||
for(int j = 0; j < 6; j++ )
|
||||
par.cameras[i].adfGeoTransform[j] = adfGeoTransform[j];
|
||||
par.cameras[i].yolo = &yolo[i];
|
||||
// par.cameras[i].yolo.init(par.net);
|
||||
// par.cameras[i].yolo.thresh = 0.25;
|
||||
// cameras[i].yolo = yolo[i];
|
||||
// cameras[i].yolo = yolo;
|
||||
|
||||
}
|
||||
pthread_t camera_task[par.n_cameras];
|
||||
for(int i=0; i<par.n_cameras; i++)
|
||||
{
|
||||
std::cout<<"creating thread\n";
|
||||
if(pthread_create(&camera_task[i], NULL, computationTask, (void*)&(par.cameras[i])))
|
||||
{
|
||||
fprintf(stderr, "error creating thread\n");
|
||||
return 1;
|
||||
}
|
||||
}
|
||||
for(int i=0; i<par.n_cameras; i++)
|
||||
{
|
||||
pthread_join(camera_task[i], NULL);
|
||||
}
|
||||
std::cout <<" free adfGeoT \n";
|
||||
free(adfGeoTransform);
|
||||
std::cout << "detection end\n";
|
||||
return 0;
|
||||
}
|
||||
@@ -1,248 +0,0 @@
|
||||
#ifndef LAYER_H
|
||||
#define LAYER_H
|
||||
|
||||
#include<iostream>
|
||||
#include "utils.h"
|
||||
#include "Network.h"
|
||||
|
||||
namespace tkDNN {
|
||||
|
||||
/**
|
||||
Data rapresentation beetween layers
|
||||
n = batch size
|
||||
c = channels
|
||||
h = heigth (lines)
|
||||
w = width (rows)
|
||||
l = lenght (3rd dimension)
|
||||
*/
|
||||
struct dataDim_t {
|
||||
|
||||
int n, c, h, w, l;
|
||||
|
||||
dataDim_t() : n(1), c(1), h(1), w(1), l(1) {};
|
||||
|
||||
dataDim_t(int _n, int _c, int _h, int _w, int _l = 1) :
|
||||
n(_n), c(_c), h(_h), w(_w), l(_l) {};
|
||||
|
||||
void print() {
|
||||
std::cout<<"Data dim: "<<n<<" "<<c<<" "<<h<<" "<<w<<" "<<l<<"\n";
|
||||
}
|
||||
|
||||
int tot() {
|
||||
return n*c*h*w*l;
|
||||
}
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
Simple layer Father class
|
||||
*/
|
||||
class Layer {
|
||||
|
||||
public:
|
||||
Layer(Network *net, dataDim_t input_dim);
|
||||
virtual ~Layer();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData) {
|
||||
std::cout<<"No infer action for this layer\n";
|
||||
return NULL;
|
||||
}
|
||||
|
||||
dataDim_t input_dim, output_dim;
|
||||
|
||||
protected:
|
||||
Network *net;
|
||||
cudnnTensorDescriptor_t srcTensorDesc, dstTensorDesc;
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
Father class of all layer that need to load trained weights
|
||||
*/
|
||||
class LayerWgs : public Layer {
|
||||
|
||||
public:
|
||||
LayerWgs(Network *net, dataDim_t input_dim,
|
||||
int inputs, int outputs, int kh, int kw, int kt,
|
||||
const char* fname_weights, const char* fname_bias);
|
||||
virtual ~LayerWgs();
|
||||
|
||||
protected:
|
||||
int inputs, outputs;
|
||||
std::string weights_path, bias_path;
|
||||
|
||||
value_type *data_h, *data_d;
|
||||
value_type *bias_h, *bias_d;
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
Dense (full interconnection) layer
|
||||
*/
|
||||
class Dense : public LayerWgs {
|
||||
|
||||
public:
|
||||
Dense(Network *net, dataDim_t in_dim, int out_ch,
|
||||
const char* fname_weights, const char* fname_bias);
|
||||
virtual ~Dense();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
value_type *dstData; //where results will be putted
|
||||
};
|
||||
|
||||
/**
|
||||
Avaible activation functions
|
||||
*/
|
||||
typedef enum {
|
||||
ACTIVATION_SIGMOID = 0,
|
||||
ACTIVATION_RELU = 1,
|
||||
ACTIVATION_TANH = 2,
|
||||
ACTIVATION_ELU = 100
|
||||
} tkdnnActivationMode_t;
|
||||
|
||||
/**
|
||||
Activation layer (it doesnt need weigths)
|
||||
*/
|
||||
class Activation : public Layer {
|
||||
|
||||
public:
|
||||
Activation(Network *net, dataDim_t input_dim, tkdnnActivationMode_t act_mode);
|
||||
virtual ~Activation();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
tkdnnActivationMode_t act_mode;
|
||||
value_type *dstData; //where results will be putted
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
Convolutional 2D layer
|
||||
*/
|
||||
class Conv2d : public LayerWgs {
|
||||
|
||||
public:
|
||||
Conv2d(Network *net, dataDim_t in_dim, int out_ch,
|
||||
int kernelH, int kernelW, int strideH, int strideW,
|
||||
const char* fname_weights, const char* fname_bias);
|
||||
virtual ~Conv2d();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
value_type *dstData; //where results will be putted
|
||||
int kernelH, kernelW, strideH, strideW;
|
||||
|
||||
cudnnFilterDescriptor_t filterDesc;
|
||||
cudnnConvolutionDescriptor_t convDesc;
|
||||
cudnnConvolutionFwdAlgo_t algo;
|
||||
cudnnTensorDescriptor_t biasTensorDesc;
|
||||
|
||||
void* workSpace;
|
||||
size_t ws_sizeInBytes;
|
||||
};
|
||||
|
||||
/**
|
||||
Convolutional 3D layer
|
||||
*/
|
||||
class Conv3d : public LayerWgs {
|
||||
|
||||
public:
|
||||
Conv3d(Network *net, dataDim_t in_dim, int out_ch,
|
||||
int kernelH, int kernelW, int kernelL,
|
||||
int strideH, int strideW, int strideL,
|
||||
const char* fname_weights, const char* fname_bias);
|
||||
virtual ~Conv3d();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
value_type *dstData; //where results will be putted
|
||||
int kernelH, kernelW, kernelL;
|
||||
int strideH, strideW, strideL;
|
||||
|
||||
cudnnFilterDescriptor_t filterDesc;
|
||||
cudnnConvolutionDescriptor_t convDesc;
|
||||
cudnnConvolutionFwdAlgo_t algo;
|
||||
cudnnTensorDescriptor_t biasTensorDesc;
|
||||
cudnnTensorDescriptor_t biasDstTensorDesc;
|
||||
|
||||
void* workSpace;
|
||||
size_t ws_sizeInBytes;
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
Flatten layer
|
||||
is actually a matrix transposition
|
||||
*/
|
||||
class Flatten : public Layer {
|
||||
|
||||
public:
|
||||
Flatten(Network *net, dataDim_t input_dim);
|
||||
virtual ~Flatten();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
value_type *dstData; //where results will be putted
|
||||
};
|
||||
|
||||
|
||||
/**
|
||||
MulAdd layer
|
||||
apply a multiplication and then an addition for each data
|
||||
*/
|
||||
class MulAdd : public Layer {
|
||||
|
||||
public:
|
||||
MulAdd(Network *net, dataDim_t input_dim, value_type mul, value_type add);
|
||||
virtual ~MulAdd();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
value_type mul, add;
|
||||
value_type *dstData, *add_vector; //where results will be putted
|
||||
};
|
||||
|
||||
|
||||
|
||||
/**
|
||||
Avaible pooling functions (padding on tkDNN is not supported)
|
||||
*/
|
||||
typedef enum {
|
||||
POOLING_MAX = 0,
|
||||
POOLING_AVERAGE = 1, // count for average includes padded values
|
||||
POOLING_AVERAGE_EXCLUDE_PADDING = 2 // count for average does not include padded values
|
||||
} tkdnnPoolingMode_t;
|
||||
|
||||
/**
|
||||
Pooling layer
|
||||
currenty supported only 2d pooing (also on 3d input)
|
||||
*/
|
||||
class Pooling : public Layer {
|
||||
|
||||
public:
|
||||
Pooling(Network *net, dataDim_t input_dim, int winH, int winW,
|
||||
int strideH, int strideW, tkdnnPoolingMode_t pool_mode);
|
||||
virtual ~Pooling();
|
||||
|
||||
virtual value_type* infer(dataDim_t &dim, value_type* srcData);
|
||||
|
||||
protected:
|
||||
|
||||
cudnnPoolingDescriptor_t poolingDesc;
|
||||
|
||||
int winH, winW;
|
||||
int strideH, strideW;
|
||||
tkdnnPoolingMode_t pool_mode;
|
||||
value_type *dstData, *tmpInputData, *tmpOutputData; //where results will be putted
|
||||
bool poolOn3d;
|
||||
};
|
||||
|
||||
}
|
||||
#endif //LAYER_H
|
||||
@@ -1,36 +0,0 @@
|
||||
#ifndef NETWORK_H
|
||||
#define NETWORK_H
|
||||
|
||||
#include "utils.h"
|
||||
|
||||
namespace tkDNN {
|
||||
|
||||
struct dataDim_t;
|
||||
class Layer;
|
||||
const int MAX_LAYERS = 256;
|
||||
|
||||
class Network {
|
||||
|
||||
public:
|
||||
Network();
|
||||
virtual ~Network();
|
||||
|
||||
/**
|
||||
Do inferece for every added layer
|
||||
*/
|
||||
value_type* infer(dataDim_t &dim, value_type* data);
|
||||
|
||||
bool addLayer(Layer *l);
|
||||
|
||||
cudnnDataType_t dataType;
|
||||
cudnnTensorFormat_t tensorFormat;
|
||||
cudnnHandle_t cudnnHandle;
|
||||
cublasHandle_t cublasHandle;
|
||||
|
||||
private:
|
||||
Layer* layers[MAX_LAYERS]; //contains layers of the net
|
||||
int num_layers; //current number of layers
|
||||
};
|
||||
|
||||
}
|
||||
#endif //NETWORK_H
|
||||
@@ -0,0 +1,44 @@
|
||||
#include <iostream>
|
||||
#include <sstream>
|
||||
#include <fstream>
|
||||
#include <iomanip>
|
||||
#include <stdlib.h>
|
||||
#include <cstring>
|
||||
#include <cstdlib>
|
||||
#include <time.h>
|
||||
#include <chrono>
|
||||
|
||||
#include "cuda.h"
|
||||
#include "cuda_runtime_api.h"
|
||||
#include <cublas_v2.h>
|
||||
#include <cudnn.h>
|
||||
|
||||
#include <opencv2/core/core.hpp>
|
||||
#include <opencv2/highgui/highgui.hpp>
|
||||
#include <opencv2/imgproc/imgproc.hpp>
|
||||
|
||||
//saliency
|
||||
#include <opencv2/core/utility.hpp>
|
||||
//#include <opencv2/saliency.hpp>
|
||||
#include <opencv2/highgui.hpp>
|
||||
|
||||
#define SAVE false
|
||||
#define SAVE_TO(name, fn, i, var) {sprintf(buf_frame_crop_name,name,fn,i);\
|
||||
cv::imwrite(buf_frame_crop_name, var);}
|
||||
|
||||
|
||||
|
||||
// cv::Mat img_threshold(cv::Mat frame_crop);
|
||||
// cv::Mat img_background(cv::Mat frame_crop);
|
||||
// cv::Mat img_dist_transform(cv::Mat frame_crop);
|
||||
// cv::Mat img_watershed(cv::Mat frame_crop);
|
||||
void image_segmentation(cv::Mat frame_crop, int frame_nbr, int i);
|
||||
void image_gradients(cv::Mat frame_crop, int frame_nbr, int i);
|
||||
void image_find_contours(cv::Mat frame_crop, int frame_nbr, int i);
|
||||
void image_saliency(cv::Mat frame_crop, int frame_nbr, int i);
|
||||
void frame_box_disparity(cv::Mat pre_frame, cv::Mat frame, std::vector <cv::Rect> pre_rois, int frame_nbr);
|
||||
void segmentation(cv::Mat pre_frame, cv::Mat frame_crop, int frame_nbr, int i, int mode);
|
||||
|
||||
//canny
|
||||
cv::Mat img_laplacian(cv::Mat frame_crop, int ret);
|
||||
cv::Mat frame_disparity(cv::Mat pre_frame, cv::Mat frame, int frame_nbr, int i, int ret);
|
||||
@@ -0,0 +1,32 @@
|
||||
#ifndef CALIBRATION_H
|
||||
#define CALIBRATION_H
|
||||
|
||||
#include "gdal.h"
|
||||
#include <gdal_priv.h>
|
||||
#include <gdal/gdal.h>
|
||||
#include "gdal/gdal_priv.h"
|
||||
#include "gdal/cpl_conv.h"
|
||||
|
||||
#include <yaml-cpp/yaml.h>
|
||||
#include <opencv2/calib3d.hpp>
|
||||
#include <opencv2/core.hpp>
|
||||
|
||||
#include <iostream>
|
||||
#include <cstring>
|
||||
|
||||
struct ObjCoords
|
||||
{
|
||||
double lat_;
|
||||
double long_;
|
||||
int class_;
|
||||
};
|
||||
|
||||
void readTiff(char *filename, double *adfGeoTransform);
|
||||
void readCameraCalibrationYaml(const std::string &cameraCalib, cv::Mat &cameraMat, cv::Mat &distCoeff);
|
||||
void pixel2coord(int x, int y, double &lat, double &lon, double *adfGeoTransform);
|
||||
void coord2pixel(double lat, double lon, int &x, int &y, double *adfGeoTransform);
|
||||
void fillMatrix(cv::Mat &H, double *matrix, bool show = false);
|
||||
void read_projection_matrix(cv::Mat &H, char *path);
|
||||
void convert_coords(std::vector<ObjCoords> &coords, int x, int y, int detected_class, cv::Mat H, double *adfGeoTransform);
|
||||
|
||||
#endif /*CALIBRATION_H*/
|
||||
@@ -0,0 +1,45 @@
|
||||
#ifndef CAMERAUTILS_H
|
||||
#define CAMERAUTILS_H
|
||||
|
||||
#include <vector>
|
||||
#include <mutex>
|
||||
#include <opencv2/core/core.hpp>
|
||||
#include "tracker.h"
|
||||
#include "Yolo3Detection.h"
|
||||
|
||||
struct Camera_t
|
||||
{
|
||||
int CAM_IDX;
|
||||
char *input;
|
||||
char *pmatrix;
|
||||
char *maskfile;
|
||||
char *cameraCalib;
|
||||
char *maskFileOrient;
|
||||
bool to_show;
|
||||
tk::dnn::Yolo3Detection *yolo;
|
||||
double adfGeoTransform[6];
|
||||
};
|
||||
|
||||
struct Frame_t
|
||||
{
|
||||
char *input;
|
||||
cv::Mat frame;
|
||||
int frame_nbr;
|
||||
// sem_vc for mainthread, videocapturethread, originalthread and disparitythread
|
||||
std::mutex sem_vc;
|
||||
};
|
||||
|
||||
struct ModFrame_t
|
||||
{
|
||||
std::vector<Tracker> trackers;
|
||||
geodetic_converter::GeodeticConverter gc;
|
||||
double adfGeoTransform[6];
|
||||
cv::Mat H;
|
||||
cv::Mat original_frame;
|
||||
tk::dnn::Yolo3Detection yolo;
|
||||
cv::Mat mask;
|
||||
// sem for mainthread, detectionthread and topviewthread
|
||||
std::mutex sem;
|
||||
};
|
||||
|
||||
#endif /*CAMERAUTILS_H*/
|
||||
@@ -0,0 +1,21 @@
|
||||
#ifndef CONFIGURATION_H
|
||||
#define CONFIGURATION_H
|
||||
|
||||
#include "cameraUtils.h"
|
||||
#include <iostream>
|
||||
#include <cstring>
|
||||
|
||||
#include <yaml-cpp/yaml.h>
|
||||
|
||||
struct Parameters_t
|
||||
{
|
||||
char *net;
|
||||
char *tiffile;
|
||||
int n_cameras;
|
||||
Camera_t *cameras;
|
||||
};
|
||||
|
||||
void readCamerasParametersYaml(const std::string &camerasParams, Parameters_t *par);
|
||||
bool read_parameters(int argc, char *argv[], Parameters_t *par);
|
||||
|
||||
#endif /*CONFIGURATION_H*/
|
||||
@@ -1,3 +0,0 @@
|
||||
#include "utils.h"
|
||||
|
||||
void activationELUForward(value_type* srcData, value_type* dstData, int size);
|
||||
@@ -0,0 +1,22 @@
|
||||
#ifndef MESSAGE_H
|
||||
#define MESSAGE_H
|
||||
|
||||
#include <iostream>
|
||||
#include <cstdlib>
|
||||
#include <ctime>
|
||||
#include <opencv2/calib3d.hpp>
|
||||
#include <opencv2/core.hpp>
|
||||
// #include <sys/socket.h> //socket
|
||||
// #include <arpa/inet.h> //inet_addr
|
||||
// #include <unistd.h> //write
|
||||
|
||||
#include "tracker.h"
|
||||
|
||||
#include "../masa_protocol/include/send.hpp"
|
||||
#include "../masa_protocol/include/serialize.hpp"
|
||||
|
||||
unsigned long long time_in_ms();
|
||||
|
||||
void addRoadUserfromTracker(const std::vector<Tracker> &trackers, Message *m, geodetic_converter::GeodeticConverter &gc, const cv::Mat &maskOrient, double *adfGeoTransform, cv::Mat H);
|
||||
|
||||
#endif /*MESSAGE_H*/
|
||||
@@ -0,0 +1,538 @@
|
||||
#ifndef LAYER_H
|
||||
#define LAYER_H
|
||||
|
||||
#include<iostream>
|
||||
#include<vector>
|
||||
#include "utils.h"
|
||||
#include "Network.h"
|
||||
|
||||
namespace tk
|
||||
{
|
||||
namespace dnn
|
||||
{
|
||||
|
||||
enum layerType_t {
|
||||
LAYER_INPUT,
|
||||
LAYER_DENSE,
|
||||
LAYER_CONV2D,
|
||||
LAYER_LSTM,
|
||||
LAYER_ACTIVATION,
|
||||
LAYER_FLATTEN,
|
||||
LAYER_MULADD,
|
||||
LAYER_POOLING,
|
||||
LAYER_SOFTMAX,
|
||||
LAYER_ROUTE,
|
||||
LAYER_REORG,
|
||||
LAYER_SHORTCUT,
|
||||
LAYER_UPSAMPLE,
|
||||
LAYER_REGION,
|
||||
LAYER_YOLO
|
||||
};
|
||||
|
||||
#define TKDNN_BN_MIN_EPSILON 1e-5
|
||||
|
||||
/**
|
||||
Simple layer Father class
|
||||
*/
|
||||
class Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Layer(Network *net);
|
||||
virtual ~Layer();
|
||||
virtual layerType_t getLayerType() = 0;
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData)
|
||||
{
|
||||
std::cout << "No infer action for this layer\n";
|
||||
return NULL;
|
||||
}
|
||||
|
||||
dataDim_t input_dim, output_dim;
|
||||
dnnType *dstData; //where results will be putted
|
||||
|
||||
std::string getLayerName()
|
||||
{
|
||||
layerType_t type = getLayerType();
|
||||
switch(type) {
|
||||
case LAYER_INPUT: return "Input";
|
||||
case LAYER_DENSE: return "Dense";
|
||||
case LAYER_CONV2D: return "Conv2d";
|
||||
case LAYER_LSTM: return "LSTM";
|
||||
case LAYER_ACTIVATION: return "Activation";
|
||||
case LAYER_FLATTEN: return "Flatten";
|
||||
case LAYER_MULADD: return "MulAdd";
|
||||
case LAYER_POOLING: return "Pooling";
|
||||
case LAYER_SOFTMAX: return "Softmax";
|
||||
case LAYER_ROUTE: return "Route";
|
||||
case LAYER_REORG: return "Reorg";
|
||||
case LAYER_SHORTCUT: return "Shortcut";
|
||||
case LAYER_UPSAMPLE: return "Upsample";
|
||||
case LAYER_REGION: return "Region";
|
||||
case LAYER_YOLO: return "Yolo";
|
||||
default: return "unknown";
|
||||
}
|
||||
}
|
||||
|
||||
protected:
|
||||
Network *net;
|
||||
cudnnTensorDescriptor_t srcTensorDesc, dstTensorDesc;
|
||||
};
|
||||
|
||||
/**
|
||||
Father class of all layer that need to load trained weights
|
||||
*/
|
||||
class LayerWgs : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
LayerWgs(Network *net, int inputs, int outputs, int kh, int kw, int kt,
|
||||
std::string fname_weights, bool batchnorm = false);
|
||||
virtual ~LayerWgs();
|
||||
|
||||
int inputs, outputs;
|
||||
std::string weights_path;
|
||||
|
||||
dnnType *data_h, *data_d;
|
||||
dnnType *bias_h, *bias_d;
|
||||
|
||||
//batchnorm
|
||||
bool batchnorm;
|
||||
dnnType *power_h;
|
||||
dnnType *scales_h, *scales_d;
|
||||
dnnType *mean_h, *mean_d;
|
||||
dnnType *variance_h, *variance_d;
|
||||
|
||||
//fp16
|
||||
__half *data16_h, *bias16_h;
|
||||
__half *data16_d, *bias16_d;
|
||||
|
||||
__half *power16_h, *power16_d;
|
||||
__half *scales16_h, *scales16_d;
|
||||
__half *mean16_h, *mean16_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
|
||||
*/
|
||||
class Dense : public LayerWgs
|
||||
{
|
||||
|
||||
public:
|
||||
Dense(Network *net, int out_ch, std::string fname_weights);
|
||||
virtual ~Dense();
|
||||
virtual layerType_t getLayerType() { return LAYER_DENSE; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
};
|
||||
|
||||
/**
|
||||
Avaible activation functions
|
||||
*/
|
||||
typedef enum
|
||||
{
|
||||
ACTIVATION_ELU = 100,
|
||||
ACTIVATION_LEAKY = 101
|
||||
} tkdnnActivationMode_t;
|
||||
|
||||
/**
|
||||
Activation layer (it doesnt need weigths)
|
||||
*/
|
||||
class Activation : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
int act_mode;
|
||||
|
||||
Activation(Network *net, int act_mode);
|
||||
virtual ~Activation();
|
||||
virtual layerType_t getLayerType() { return LAYER_ACTIVATION; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
protected:
|
||||
cudnnActivationDescriptor_t activDesc;
|
||||
};
|
||||
|
||||
/**
|
||||
Convolutional 2D layer
|
||||
|
||||
WEIGHTS shape: OUTCH, INCH, KH, KW ...
|
||||
BIAS shape: OUTCH
|
||||
|
||||
with BATCHNORM:
|
||||
scales: OUTCH
|
||||
means: OUTCH
|
||||
variance: OUTCH
|
||||
*/
|
||||
class Conv2d : public LayerWgs
|
||||
{
|
||||
|
||||
public:
|
||||
Conv2d( Network *net, int out_ch, int kernelH, int kernelW,
|
||||
int strideH, int strideW, int paddingH, int paddingW,
|
||||
std::string fname_weights, bool batchnorm = false);
|
||||
virtual ~Conv2d();
|
||||
virtual layerType_t getLayerType() { return LAYER_CONV2D; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
int kernelH, kernelW, strideH, strideW, paddingH, paddingW;
|
||||
|
||||
protected:
|
||||
cudnnFilterDescriptor_t filterDesc;
|
||||
cudnnConvolutionDescriptor_t convDesc;
|
||||
cudnnConvolutionFwdAlgo_t algo;
|
||||
cudnnTensorDescriptor_t biasTensorDesc;
|
||||
|
||||
void *workSpace;
|
||||
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
|
||||
is actually a matrix transposition
|
||||
*/
|
||||
class Flatten : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Flatten(Network *net);
|
||||
virtual ~Flatten();
|
||||
virtual layerType_t getLayerType() { return LAYER_FLATTEN; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
};
|
||||
|
||||
/**
|
||||
MulAdd layer
|
||||
apply a multiplication and then an addition for each data
|
||||
*/
|
||||
class MulAdd : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
MulAdd(Network *net, dnnType mul, dnnType add);
|
||||
virtual ~MulAdd();
|
||||
virtual layerType_t getLayerType() { return LAYER_MULADD; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
protected:
|
||||
dnnType mul, add;
|
||||
dnnType *add_vector;
|
||||
};
|
||||
|
||||
/**
|
||||
Avaible pooling functions (padding on tkDNN is not supported)
|
||||
*/
|
||||
typedef enum
|
||||
{
|
||||
POOLING_MAX = 0,
|
||||
POOLING_AVERAGE = 1, // count for average includes padded values
|
||||
POOLING_AVERAGE_EXCLUDE_PADDING = 2 // count for average does not include padded values
|
||||
} tkdnnPoolingMode_t;
|
||||
|
||||
/**
|
||||
Pooling layer
|
||||
currenty supported only 2d pooing (also on 3d input)
|
||||
*/
|
||||
class Pooling : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
int winH, winW;
|
||||
int strideH, strideW;
|
||||
int paddingH, paddingW;
|
||||
|
||||
Pooling(Network *net, int winH, int winW,
|
||||
int strideH, int strideW, tkdnnPoolingMode_t pool_mode);
|
||||
virtual ~Pooling();
|
||||
virtual layerType_t getLayerType() { return LAYER_POOLING; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
protected:
|
||||
cudnnPoolingDescriptor_t poolingDesc;
|
||||
tkdnnPoolingMode_t pool_mode;
|
||||
dnnType *tmpInputData, *tmpOutputData;
|
||||
bool poolOn3d;
|
||||
};
|
||||
|
||||
/**
|
||||
Softmax layer
|
||||
*/
|
||||
class Softmax : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Softmax(Network *net);
|
||||
virtual ~Softmax();
|
||||
virtual layerType_t getLayerType() { return LAYER_SOFTMAX; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
};
|
||||
|
||||
/**
|
||||
Route layer
|
||||
Merge a list of layers
|
||||
*/
|
||||
class Route : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Route(Network *net, Layer **layers, int layers_n);
|
||||
virtual ~Route();
|
||||
virtual layerType_t getLayerType() { return LAYER_ROUTE; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
public:
|
||||
Layer **layers; //ids of layers to be merged
|
||||
int layers_n; //number of layers
|
||||
};
|
||||
|
||||
/**
|
||||
Reorg layer
|
||||
Mantain same dimension but change C*H*W distribution
|
||||
*/
|
||||
class Reorg : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Reorg(Network *net, int stride);
|
||||
virtual ~Reorg();
|
||||
virtual layerType_t getLayerType() { return LAYER_REORG; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
int stride;
|
||||
};
|
||||
|
||||
/**
|
||||
Shortcut layer
|
||||
sum with stride another layer
|
||||
*/
|
||||
class Shortcut : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Shortcut(Network *net, Layer *backLayer);
|
||||
virtual ~Shortcut();
|
||||
virtual layerType_t getLayerType() { return LAYER_SHORTCUT; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
public:
|
||||
Layer *backLayer;
|
||||
};
|
||||
|
||||
/**
|
||||
Upsample layer
|
||||
Mantain same dimension but change C*H*W distribution
|
||||
*/
|
||||
class Upsample : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Upsample(Network *net, int stride);
|
||||
virtual ~Upsample();
|
||||
virtual layerType_t getLayerType() { return LAYER_UPSAMPLE; };
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
|
||||
int stride;
|
||||
bool reverse;
|
||||
};
|
||||
|
||||
struct box
|
||||
{
|
||||
int cl;
|
||||
float x, y, w, h;
|
||||
float prob;
|
||||
};
|
||||
struct sortable_bbox
|
||||
{
|
||||
int index;
|
||||
int cl;
|
||||
float **probs;
|
||||
};
|
||||
|
||||
/**
|
||||
Yolo3 layer
|
||||
*/
|
||||
class Yolo : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
struct box
|
||||
{
|
||||
float x, y, w, h;
|
||||
};
|
||||
|
||||
struct detection
|
||||
{
|
||||
Yolo::box bbox;
|
||||
int classes;
|
||||
float *prob;
|
||||
float *mask;
|
||||
float objectness;
|
||||
int sort_class;
|
||||
};
|
||||
|
||||
Yolo(Network *net, int classes, int num, std::string fname_weights);
|
||||
virtual ~Yolo();
|
||||
virtual layerType_t getLayerType() { return LAYER_YOLO; };
|
||||
|
||||
int classes, num;
|
||||
dnnType *mask_h, *mask_d; //anchors
|
||||
dnnType *bias_h, *bias_d; //anchors
|
||||
std::vector<std::string> classesNames;
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
int computeDetections(Yolo::detection *dets, int &ndets, int netw, int neth, float thresh);
|
||||
|
||||
dnnType *predictions;
|
||||
|
||||
static const int MAX_DETECTIONS = 256;
|
||||
static Yolo::detection *allocateDetections(int nboxes, int classes);
|
||||
static void mergeDetections(Yolo::detection *dets, int ndets, int classes);
|
||||
};
|
||||
|
||||
/**
|
||||
Region layer
|
||||
*/
|
||||
class Region : public Layer
|
||||
{
|
||||
|
||||
public:
|
||||
Region(Network *net, int classes, int coords, int num);
|
||||
virtual ~Region();
|
||||
virtual layerType_t getLayerType() { return LAYER_REGION; };
|
||||
|
||||
int classes, coords, num;
|
||||
|
||||
virtual dnnType *infer(dataDim_t &dim, dnnType *srcData);
|
||||
};
|
||||
|
||||
class RegionInterpret
|
||||
{
|
||||
|
||||
public:
|
||||
RegionInterpret(dataDim_t input_dim, dataDim_t output_dim,
|
||||
int classes, int coords, int num, float thresh, std::string fname_weights);
|
||||
~RegionInterpret();
|
||||
|
||||
dataDim_t input_dim, output_dim;
|
||||
dnnType *bias_h, *bias_d; //anchors
|
||||
int classes, coords, num;
|
||||
float thresh;
|
||||
|
||||
box *boxes;
|
||||
float **probs;
|
||||
sortable_bbox *s;
|
||||
box res_boxes[256];
|
||||
int res_boxes_n;
|
||||
|
||||
box get_region_box(float *x, float *biases, int n, int index, int i, int j, int w, int h, int stride);
|
||||
void get_region_boxes(float *input, int w, int h, int netw, int neth, float thresh,
|
||||
float **probs, box *boxes, int only_objectness,
|
||||
int *map, float tree_thresh, int relative);
|
||||
void correct_region_boxes(box *boxes, int n, int w, int h, int netw, int neth, int relative);
|
||||
void interpretData(dnnType *data_h, int imageW = 0, int imageH = 0);
|
||||
void showImageResult(dnnType *input_h);
|
||||
|
||||
static float box_iou(box a, box b);
|
||||
};
|
||||
|
||||
} // namespace dnn
|
||||
} // namespace tk
|
||||
#endif //LAYER_H
|
||||
@@ -0,0 +1,73 @@
|
||||
#ifndef NETWORK_H
|
||||
#define NETWORK_H
|
||||
|
||||
#include "utils.h"
|
||||
|
||||
namespace tk
|
||||
{
|
||||
namespace dnn
|
||||
{
|
||||
|
||||
/**
|
||||
Data rapresentation beetween layers
|
||||
n = batch size
|
||||
c = channels
|
||||
h = heigth (lines)
|
||||
w = width (rows)
|
||||
l = lenght (3rd dimension)
|
||||
*/
|
||||
struct dataDim_t
|
||||
{
|
||||
|
||||
int n, c, h, w, l;
|
||||
|
||||
dataDim_t() : n(1), c(1), h(1), w(1), l(1){};
|
||||
|
||||
dataDim_t(int _n, int _c, int _h, int _w, int _l = 1) : n(_n), c(_c), h(_h), w(_w), l(_l){};
|
||||
|
||||
void print()
|
||||
{
|
||||
std::cout << "Data dim: " << n << " " << c << " " << h << " " << w << " " << l << "\n";
|
||||
}
|
||||
|
||||
int tot()
|
||||
{
|
||||
return n * c * h * w * l;
|
||||
}
|
||||
};
|
||||
|
||||
class Layer;
|
||||
const int MAX_LAYERS = 256;
|
||||
|
||||
class Network
|
||||
{
|
||||
|
||||
public:
|
||||
Network(dataDim_t input_dim);
|
||||
virtual ~Network();
|
||||
|
||||
/**
|
||||
Do inferece for every added layer
|
||||
*/
|
||||
dnnType *infer(dataDim_t &dim, dnnType *data);
|
||||
|
||||
bool addLayer(Layer *l);
|
||||
void print();
|
||||
|
||||
cudnnDataType_t dataType;
|
||||
cudnnTensorFormat_t tensorFormat;
|
||||
cudnnHandle_t cudnnHandle;
|
||||
cublasHandle_t cublasHandle;
|
||||
|
||||
Layer *layers[MAX_LAYERS]; //contains layers of the net
|
||||
int num_layers; //current number of layers
|
||||
|
||||
dataDim_t input_dim;
|
||||
dataDim_t getOutputDim();
|
||||
|
||||
bool fp16, dla;
|
||||
};
|
||||
|
||||
} // namespace dnn
|
||||
} // namespace tk
|
||||
#endif //NETWORK_H
|
||||
@@ -0,0 +1,99 @@
|
||||
#ifndef NETWORKRT_H
|
||||
#define NETWORKRT_H
|
||||
|
||||
#include <string.h> // memcpy
|
||||
#include "utils.h"
|
||||
#include "Network.h"
|
||||
#include "Layer.h"
|
||||
#include "NvInfer.h"
|
||||
|
||||
namespace tk
|
||||
{
|
||||
namespace dnn
|
||||
{
|
||||
|
||||
template <typename T>
|
||||
void writeBUF(char *&buffer, const T &val)
|
||||
{
|
||||
*reinterpret_cast<T *>(buffer) = val;
|
||||
buffer += sizeof(T);
|
||||
}
|
||||
|
||||
template <typename T>
|
||||
T readBUF(const char *&buffer)
|
||||
{
|
||||
T val = *reinterpret_cast<const T *>(buffer);
|
||||
buffer += sizeof(T);
|
||||
return val;
|
||||
}
|
||||
|
||||
using namespace nvinfer1;
|
||||
#include "pluginsRT/ActivationLeakyRT.h"
|
||||
#include "pluginsRT/ReorgRT.h"
|
||||
#include "pluginsRT/RegionRT.h"
|
||||
//#include "pluginsRT/RouteRT.h"
|
||||
#include "pluginsRT/ShortcutRT.h"
|
||||
#include "pluginsRT/YoloRT.h"
|
||||
#include "pluginsRT/UpsampleRT.h"
|
||||
//#include "pluginsRT/Int8Calibrator.h"
|
||||
|
||||
class PluginFactory : IPluginFactory
|
||||
{
|
||||
public:
|
||||
YoloRT *yolos[16];
|
||||
int n_yolos;
|
||||
|
||||
virtual IPlugin *createPlugin(const char *layerName, const void *serialData, size_t serialLength);
|
||||
};
|
||||
|
||||
class NetworkRT
|
||||
{
|
||||
|
||||
public:
|
||||
nvinfer1::DataType dtRT;
|
||||
nvinfer1::IBuilder *builderRT;
|
||||
nvinfer1::IRuntime *runtimeRT;
|
||||
nvinfer1::INetworkDefinition *networkRT;
|
||||
|
||||
nvinfer1::ICudaEngine *engineRT;
|
||||
nvinfer1::IExecutionContext *contextRT;
|
||||
|
||||
const static int MAX_BUFFERS_RT = 10;
|
||||
void *buffersRT[MAX_BUFFERS_RT];
|
||||
int buf_input_idx, buf_output_idx;
|
||||
|
||||
dataDim_t input_dim, output_dim;
|
||||
dnnType *output;
|
||||
cudaStream_t stream;
|
||||
|
||||
PluginFactory *pluginFactory;
|
||||
|
||||
NetworkRT(Network *net, const char *name);
|
||||
virtual ~NetworkRT();
|
||||
|
||||
/**
|
||||
Do inferece
|
||||
*/
|
||||
dnnType *infer(dataDim_t &dim, dnnType *data);
|
||||
void enqueue();
|
||||
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Layer *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Conv2d *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Activation *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Dense *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Pooling *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Softmax *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Route *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Reorg *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Region *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Shortcut *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Yolo *l);
|
||||
nvinfer1::ILayer *convert_layer(nvinfer1::ITensor *input, Upsample *l);
|
||||
|
||||
bool serialize(const char *filename);
|
||||
bool deserialize(const char *filename);
|
||||
};
|
||||
|
||||
} // namespace dnn
|
||||
} // namespace tk
|
||||
#endif //NETWORKRT_H
|
||||
@@ -0,0 +1,78 @@
|
||||
#ifndef YOLO3DDETECTION_H
|
||||
#define YOLO3DDETECTION_H
|
||||
|
||||
#include <iostream>
|
||||
#include <signal.h>
|
||||
#include <stdlib.h> /* srand, rand */
|
||||
#include <unistd.h>
|
||||
#include <mutex>
|
||||
#include "utils.h"
|
||||
|
||||
#include <opencv2/core/core.hpp>
|
||||
#include <opencv2/highgui/highgui.hpp>
|
||||
#include <opencv2/imgproc/imgproc.hpp>
|
||||
|
||||
#include "tkdnn.h"
|
||||
|
||||
namespace tk
|
||||
{
|
||||
namespace dnn
|
||||
{
|
||||
|
||||
/**
|
||||
*
|
||||
* @author Francesco Gatti
|
||||
*/
|
||||
class Yolo3Detection
|
||||
{
|
||||
|
||||
private:
|
||||
tk::dnn::NetworkRT *netRT = nullptr;
|
||||
tk::dnn::Yolo *yolo[3];
|
||||
dnnType *input, *input_d;
|
||||
|
||||
int ndets = 0;
|
||||
tk::dnn::Yolo::detection *dets = nullptr;
|
||||
|
||||
cv::Mat imageF;
|
||||
cv::Mat bgr[3];
|
||||
|
||||
public:
|
||||
int classes = 0;
|
||||
int num = 0;
|
||||
float thresh = 0.3;
|
||||
cv::Scalar colors[256];
|
||||
|
||||
// this is filled with results
|
||||
std::vector<tk::dnn::box> detected;
|
||||
|
||||
// keep track of inference times (ms)
|
||||
std::vector<double> stats;
|
||||
|
||||
Yolo3Detection() {}
|
||||
|
||||
virtual ~Yolo3Detection() {}
|
||||
|
||||
/**
|
||||
* Method used to inizialize the class
|
||||
*
|
||||
* @return Success of the initialization
|
||||
*/
|
||||
bool init(std::string tensor_path);
|
||||
void addBorders(cv::Mat &imageORIG, cv::Mat &imageWBorders, int &top, int &left);
|
||||
void update(cv::Mat &frame);
|
||||
|
||||
|
||||
tk::dnn::Yolo* getYoloLayer(int n=0) {
|
||||
if(n<3)
|
||||
return yolo[n];
|
||||
else
|
||||
return nullptr;
|
||||
}
|
||||
|
||||
};
|
||||
|
||||
} // namespace dnn
|
||||
} // namespace tk
|
||||
|
||||
#endif /*YOLO3DDETECTION_H*/
|
||||
@@ -0,0 +1,27 @@
|
||||
#ifndef KERNELS_H
|
||||
#define KERNELS_H
|
||||
|
||||
#include "utils.h"
|
||||
|
||||
void activationELUForward(dnnType* srcData, dnnType* dstData, int size, cudaStream_t stream = cudaStream_t(0));
|
||||
void activationLEAKYForward(dnnType* srcData, dnnType* dstData, int size, cudaStream_t stream = cudaStream_t(0));
|
||||
void activationLOGISTICForward(dnnType* srcData, dnnType* dstData, int size, cudaStream_t stream = cudaStream_t(0));
|
||||
|
||||
void fill(dnnType* data, int size, dnnType val, cudaStream_t stream = cudaStream_t(0));
|
||||
|
||||
void reorgForward( dnnType* srcData, dnnType* dstData,
|
||||
int n, int c, int h, int w, int stride, cudaStream_t stream = cudaStream_t(0));
|
||||
void softmaxForward(float *input, int n, int batch, int batch_offset,
|
||||
int groups, int group_offset, int stride, float temp, float *output, cudaStream_t stream = cudaStream_t(0));
|
||||
|
||||
|
||||
void shortcutForward(dnnType* srcData, dnnType* dstData, int n1, int c1, int h1, int w1, int s1,
|
||||
int n2, int c2, int h2, int w2, int s2,
|
||||
cudaStream_t stream = cudaStream_t(0));
|
||||
|
||||
void upsampleForward(dnnType* srcData, dnnType* dstData,
|
||||
int n, int c, int h, int w, int s, int forward, float scale,
|
||||
cudaStream_t stream = cudaStream_t(0));
|
||||
|
||||
void float2half(float* srcData, __half* dstData, int size, const cudaStream_t stream = cudaStream_t(0));
|
||||
#endif //KERNELS_H
|
||||
@@ -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;
|
||||
@@ -0,0 +1,60 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class ActivationLeakyRT : public IPlugin {
|
||||
|
||||
public:
|
||||
ActivationLeakyRT() {
|
||||
|
||||
|
||||
}
|
||||
|
||||
~ActivationLeakyRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return inputs[0];
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
size = 1;
|
||||
for(int i=0; i<outputDims[0].nbDims; i++)
|
||||
size *= outputDims[0].d[i];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
activationLEAKYForward((dnnType*)reinterpret_cast<const dnnType*>(inputs[0]),
|
||||
reinterpret_cast<dnnType*>(outputs[0]), size, stream);
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 1*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, size);
|
||||
}
|
||||
|
||||
int size;
|
||||
};
|
||||
@@ -0,0 +1,168 @@
|
||||
#include <vector>
|
||||
#include <assert.h>
|
||||
#include <algorithm>
|
||||
#include <iterator>
|
||||
|
||||
#include "NvInfer.h"
|
||||
|
||||
class BatchStream
|
||||
{
|
||||
public:
|
||||
BatchStream(tk::dnn::dataDim_t dim, int batchSize, int maxBatches)
|
||||
{
|
||||
mBatchSize = batchSize;
|
||||
mMaxBatches = maxBatches;
|
||||
mDims = nvinfer1::DimsNCHW{ dim.n, dim.c, dim.h, dim.w };
|
||||
mImageSize = mDims.c()*mDims.h()*mDims.w();
|
||||
mBatch.resize(mBatchSize*mImageSize, 0);
|
||||
mLabels.resize(mBatchSize, 0);
|
||||
mFileBatch.resize(mDims.n()*mImageSize, 0);
|
||||
mFileLabels.resize(mDims.n(), 0);
|
||||
reset(0);
|
||||
}
|
||||
|
||||
void reset(int firstBatch)
|
||||
{
|
||||
mBatchCount = 0;
|
||||
mFileCount = 0;
|
||||
mFileBatchPos = mDims.n();
|
||||
skip(firstBatch);
|
||||
}
|
||||
|
||||
bool next()
|
||||
{
|
||||
std::cout<<"Next batch: "<<mBatchCount<<" of "<<mMaxBatches<<"\n";
|
||||
if (mBatchCount == mMaxBatches)
|
||||
return false;
|
||||
|
||||
for (int csize = 1, batchPos = 0; batchPos < mBatchSize; batchPos += csize, mFileBatchPos += csize)
|
||||
{
|
||||
assert(mFileBatchPos > 0 && mFileBatchPos <= mDims.n());
|
||||
if (mFileBatchPos == mDims.n() && !update())
|
||||
return false;
|
||||
|
||||
// copy the smaller of: elements left to fulfill the request, or elements left in the file buffer.
|
||||
csize = std::min(mBatchSize - batchPos, mDims.n() - mFileBatchPos);
|
||||
std::copy_n(getFileBatch() + mFileBatchPos * mImageSize, csize * mImageSize, getBatch() + batchPos * mImageSize);
|
||||
std::copy_n(getFileLabels() + mFileBatchPos, csize, getLabels() + batchPos);
|
||||
}
|
||||
mBatchCount++;
|
||||
return true;
|
||||
}
|
||||
|
||||
void skip(int skipCount)
|
||||
{
|
||||
if (mBatchSize >= mDims.n() && mBatchSize%mDims.n() == 0 && mFileBatchPos == mDims.n())
|
||||
{
|
||||
mFileCount += skipCount * mBatchSize / mDims.n();
|
||||
std::cout<<mFileCount<<"\n";
|
||||
return;
|
||||
}
|
||||
|
||||
int x = mBatchCount;
|
||||
for (int i = 0; i < skipCount; i++)
|
||||
next();
|
||||
mBatchCount = x;
|
||||
}
|
||||
|
||||
float *getBatch() { return &mBatch[0]; }
|
||||
float *getLabels() { return &mLabels[0]; }
|
||||
int getBatchesRead() const { return mBatchCount; }
|
||||
int getBatchSize() const { return mBatchSize; }
|
||||
nvinfer1::DimsNCHW getDims() const { return mDims; }
|
||||
private:
|
||||
float* getFileBatch() { return &mFileBatch[0]; }
|
||||
float* getFileLabels() { return &mFileLabels[0]; }
|
||||
|
||||
bool update()
|
||||
{
|
||||
std::string inputFileName = std::string("calibBatches/batch") + std::to_string(mFileCount++);
|
||||
FILE * file = fopen(inputFileName.c_str(), "rb");
|
||||
if (!file) {
|
||||
FatalError("cant open batch calib file: " + inputFileName);
|
||||
return false;
|
||||
}
|
||||
|
||||
size_t readInputCount = fread(getFileBatch(), sizeof(float), mDims.n()*mImageSize, file);
|
||||
size_t readLabelCount = fread(getFileLabels(), sizeof(float), mDims.n(), file);;
|
||||
assert(readInputCount == size_t(mDims.n()*mImageSize) && readLabelCount == size_t(mDims.n()));
|
||||
|
||||
fclose(file);
|
||||
mFileBatchPos = 0;
|
||||
return true;
|
||||
}
|
||||
|
||||
int mBatchSize{ 0 };
|
||||
int mMaxBatches{ 0 };
|
||||
int mBatchCount{ 0 };
|
||||
|
||||
int mFileCount{ 0 }, mFileBatchPos{ 0 };
|
||||
int mImageSize{ 0 };
|
||||
|
||||
nvinfer1::DimsNCHW mDims;
|
||||
std::vector<float> mBatch;
|
||||
std::vector<float> mLabels;
|
||||
std::vector<float> mFileBatch;
|
||||
std::vector<float> mFileLabels;
|
||||
};
|
||||
|
||||
|
||||
|
||||
class Int8EntropyCalibrator : public IInt8EntropyCalibrator
|
||||
{
|
||||
public:
|
||||
Int8EntropyCalibrator(BatchStream& stream, int firstBatch, bool readCache = true)
|
||||
: mStream(stream), mReadCache(readCache)
|
||||
{
|
||||
DimsNCHW dims = mStream.getDims();
|
||||
mInputCount = mStream.getBatchSize() * dims.c() * dims.h() * dims.w();
|
||||
checkCuda(cudaMalloc(&mDeviceInput, mInputCount * sizeof(float)));
|
||||
mStream.reset(firstBatch);
|
||||
}
|
||||
|
||||
virtual ~Int8EntropyCalibrator()
|
||||
{
|
||||
checkCuda(cudaFree(mDeviceInput));
|
||||
}
|
||||
|
||||
int getBatchSize() const override { return mStream.getBatchSize(); }
|
||||
|
||||
bool getBatch(void* bindings[], const char* names[], int nbBindings) override
|
||||
{
|
||||
std::cout<<"CALIB request batch\n";
|
||||
if (!mStream.next())
|
||||
return false;
|
||||
|
||||
checkCuda(cudaMemcpy(mDeviceInput, mStream.getBatch(), mInputCount * sizeof(float), cudaMemcpyHostToDevice));
|
||||
bindings[0] = mDeviceInput;
|
||||
return true;
|
||||
}
|
||||
|
||||
const void* readCalibrationCache(size_t& length) override
|
||||
{
|
||||
mCalibrationCache.clear();
|
||||
std::ifstream input("table.calib", std::ios::binary);
|
||||
input >> std::noskipws;
|
||||
|
||||
FatalError("rewrite different");
|
||||
//if (mReadCache && input.good())
|
||||
// std::copy(std::istream_iterator<char>(input), std::istream_iterator<char>(), std::back_inserter(mCalibrationCache));
|
||||
|
||||
length = mCalibrationCache.size();
|
||||
return length ? &mCalibrationCache[0] : nullptr;
|
||||
}
|
||||
|
||||
void writeCalibrationCache(const void* cache, size_t length) override
|
||||
{
|
||||
std::ofstream output("table.calib", std::ios::binary);
|
||||
output.write(reinterpret_cast<const char*>(cache), length);
|
||||
}
|
||||
|
||||
private:
|
||||
BatchStream mStream;
|
||||
bool mReadCache{ true };
|
||||
|
||||
size_t mInputCount;
|
||||
void* mDeviceInput{ nullptr };
|
||||
std::vector<char> mCalibrationCache;
|
||||
};
|
||||
@@ -0,0 +1,94 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class RegionRT : public IPlugin {
|
||||
|
||||
public:
|
||||
RegionRT(int classes, int coords, int num) {
|
||||
|
||||
this->classes = classes;
|
||||
this->coords = coords;
|
||||
this->num = num;
|
||||
}
|
||||
|
||||
~RegionRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return inputs[0];
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
c = inputDims[0].d[0];
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
dnnType *srcData = (dnnType*)reinterpret_cast<const dnnType*>(inputs[0]);
|
||||
dnnType *dstData = reinterpret_cast<dnnType*>(outputs[0]);
|
||||
|
||||
checkCuda( cudaMemcpyAsync(dstData, srcData, batchSize*c*h*w*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream));
|
||||
|
||||
for (int b = 0; b < batchSize; ++b){
|
||||
for(int n = 0; n < num; ++n){
|
||||
int index = entry_index(b, n*w*h, 0, batchSize);
|
||||
activationLOGISTICForward(srcData + index, dstData + index, 2*w*h, stream);
|
||||
|
||||
index = entry_index(b, n*w*h, coords, batchSize);
|
||||
activationLOGISTICForward(srcData + index, dstData + index, w*h, stream);
|
||||
}
|
||||
}
|
||||
|
||||
//softmax start
|
||||
int index = entry_index(0, 0, coords + 1, batchSize);
|
||||
softmaxForward( srcData + index, classes, batchSize*num,
|
||||
(batchSize*c*h*w)/num,
|
||||
w*h, 1, w*h, 1, dstData + index, stream);
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 6*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, classes);
|
||||
tk::dnn::writeBUF(buf, coords);
|
||||
tk::dnn::writeBUF(buf, num);
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
}
|
||||
|
||||
int c, h, w;
|
||||
int classes, coords, num;
|
||||
|
||||
int entry_index(int batch, int location, int entry, int batchSize) {
|
||||
int n = location / (w*h);
|
||||
int loc = location % (w*h);
|
||||
return batch*c*h*w*batchSize + n*w*h*(coords+classes+1) + entry*w*h + loc;
|
||||
}
|
||||
|
||||
};
|
||||
@@ -0,0 +1,63 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class ReorgRT : public IPlugin {
|
||||
|
||||
public:
|
||||
ReorgRT(int stride) {
|
||||
this->stride = stride;
|
||||
}
|
||||
|
||||
~ReorgRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return DimsCHW{inputs[0].d[0]*stride*stride, inputs[0].d[1]/stride, inputs[0].d[2]/stride};
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
c = inputDims[0].d[0];
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
reorgForward((dnnType*)reinterpret_cast<const dnnType*>(inputs[0]),
|
||||
reinterpret_cast<dnnType*>(outputs[0]),
|
||||
batchSize, c, h, w, stride, stream);
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 4*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, stride);
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
}
|
||||
|
||||
int c, h, w, stride;
|
||||
};
|
||||
@@ -0,0 +1,82 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class RouteRT : public IPlugin {
|
||||
|
||||
public:
|
||||
RouteRT() {
|
||||
}
|
||||
|
||||
~RouteRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
int out_c = 0;
|
||||
for(int i=0; i<nbInputDims; i++) out_c += inputs[i].d[0];
|
||||
return DimsCHW{out_c, inputs[0].d[1], inputs[0].d[2]};
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
in = nbInputs;
|
||||
c = 0;
|
||||
for(int i=0; i<nbInputs; i++) {
|
||||
c_in[i] = inputDims[i].d[0];
|
||||
c += inputDims[i].d[0];
|
||||
}
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
dnnType *dstData = reinterpret_cast<dnnType*>(outputs[0]);
|
||||
|
||||
int offset = 0;
|
||||
for(int i=0; i<in; i++) {
|
||||
dnnType *input = (dnnType*)reinterpret_cast<const dnnType*>(inputs[i]);
|
||||
int in_dim = c_in[i]*h*w;
|
||||
checkCuda( cudaMemcpyAsync(dstData + offset, input, in_dim*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream) );
|
||||
offset += in_dim;
|
||||
}
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return (4+MAX_INPUTS)*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, in);
|
||||
for(int i=0; i<MAX_INPUTS; i++)
|
||||
tk::dnn::writeBUF(buf, c_in[i]);
|
||||
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
}
|
||||
|
||||
static const int MAX_INPUTS = 4;
|
||||
int in;
|
||||
int c_in[MAX_INPUTS];
|
||||
int c, h, w;
|
||||
};
|
||||
@@ -0,0 +1,65 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class ShortcutRT : public IPlugin {
|
||||
|
||||
public:
|
||||
ShortcutRT() {
|
||||
}
|
||||
|
||||
~ShortcutRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return DimsCHW{inputs[0].d[0], inputs[0].d[1], inputs[0].d[2]};
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
c = inputDims[0].d[0];
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
dnnType *srcData = (dnnType*)reinterpret_cast<const dnnType*>(inputs[0]);
|
||||
dnnType *srcDataBack = (dnnType*)reinterpret_cast<const dnnType*>(inputs[1]);
|
||||
dnnType *dstData = reinterpret_cast<dnnType*>(outputs[0]);
|
||||
|
||||
checkCuda( cudaMemcpyAsync(dstData, srcData, batchSize*c*h*w*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream));
|
||||
shortcutForward(srcDataBack, dstData, batchSize, c, h, w, 1, batchSize, c, h, w, 1, stream);
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 3*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
}
|
||||
|
||||
int c, h, w;
|
||||
};
|
||||
@@ -0,0 +1,65 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
class UpsampleRT : public IPlugin {
|
||||
|
||||
public:
|
||||
UpsampleRT(int stride) {
|
||||
this->stride = stride;
|
||||
}
|
||||
|
||||
~UpsampleRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return DimsCHW(inputs[0].d[0], inputs[0].d[1]*stride, inputs[0].d[2]*stride);
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
c = inputDims[0].d[0];
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
dnnType *srcData = (dnnType*)reinterpret_cast<const dnnType*>(inputs[0]);
|
||||
dnnType *dstData = reinterpret_cast<dnnType*>(outputs[0]);
|
||||
|
||||
fill(dstData, batchSize*c*h*w*stride*stride, 0.0, stream);
|
||||
upsampleForward(srcData, dstData, batchSize, c, h, w, stride, 1, 1, stream);
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 4*sizeof(int);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, stride);
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
}
|
||||
|
||||
int c, h, w, stride;
|
||||
};
|
||||
@@ -0,0 +1,116 @@
|
||||
#include<cassert>
|
||||
#include "../kernels.h"
|
||||
|
||||
#define YOLORT_CLASSNAME_W 256
|
||||
|
||||
class YoloRT : public IPlugin {
|
||||
|
||||
|
||||
|
||||
public:
|
||||
YoloRT(int classes, int num, tk::dnn::Yolo *yolo = nullptr) {
|
||||
|
||||
this->classes = classes;
|
||||
this->num = num;
|
||||
|
||||
mask = new dnnType[num];
|
||||
bias = new dnnType[num*3*2];
|
||||
if(yolo != nullptr) {
|
||||
memcpy(mask, yolo->mask_h, sizeof(dnnType)*num);
|
||||
memcpy(bias, yolo->bias_h, sizeof(dnnType)*num*3*2);
|
||||
classesNames = yolo->classesNames;
|
||||
}
|
||||
}
|
||||
|
||||
~YoloRT(){
|
||||
|
||||
}
|
||||
|
||||
int getNbOutputs() const override {
|
||||
return 1;
|
||||
}
|
||||
|
||||
Dims getOutputDimensions(int index, const Dims* inputs, int nbInputDims) override {
|
||||
return inputs[0];
|
||||
}
|
||||
|
||||
void configure(const Dims* inputDims, int nbInputs, const Dims* outputDims, int nbOutputs, int maxBatchSize) override {
|
||||
c = inputDims[0].d[0];
|
||||
h = inputDims[0].d[1];
|
||||
w = inputDims[0].d[2];
|
||||
}
|
||||
|
||||
int initialize() override {
|
||||
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual void terminate() override {
|
||||
}
|
||||
|
||||
virtual size_t getWorkspaceSize(int maxBatchSize) const override {
|
||||
return 0;
|
||||
}
|
||||
|
||||
virtual int enqueue(int batchSize, const void*const * inputs, void** outputs, void* workspace, cudaStream_t stream) override {
|
||||
|
||||
dnnType *srcData = (dnnType*)reinterpret_cast<const dnnType*>(inputs[0]);
|
||||
dnnType *dstData = reinterpret_cast<dnnType*>(outputs[0]);
|
||||
|
||||
checkCuda( cudaMemcpyAsync(dstData, srcData, batchSize*c*h*w*sizeof(dnnType), cudaMemcpyDeviceToDevice, stream));
|
||||
|
||||
for (int b = 0; b < batchSize; ++b){
|
||||
for(int n = 0; n < num; ++n){
|
||||
int index = entry_index(b, n*w*h, 0, batchSize);
|
||||
activationLOGISTICForward(srcData + index, dstData + index, 2*w*h, stream);
|
||||
|
||||
index = entry_index(b, n*w*h, 4, batchSize);
|
||||
activationLOGISTICForward(srcData + index, dstData + index, (1+classes)*w*h, stream);
|
||||
}
|
||||
}
|
||||
|
||||
//std::cout<<"YOLO END\n";
|
||||
return 0;
|
||||
}
|
||||
|
||||
|
||||
virtual size_t getSerializationSize() override {
|
||||
return 5*sizeof(int) + num*sizeof(dnnType) + num*3*2*sizeof(dnnType) + YOLORT_CLASSNAME_W*classes*sizeof(char);
|
||||
}
|
||||
|
||||
virtual void serialize(void* buffer) override {
|
||||
char *buf = reinterpret_cast<char*>(buffer);
|
||||
tk::dnn::writeBUF(buf, classes);
|
||||
tk::dnn::writeBUF(buf, num);
|
||||
tk::dnn::writeBUF(buf, c);
|
||||
tk::dnn::writeBUF(buf, h);
|
||||
tk::dnn::writeBUF(buf, w);
|
||||
for(int i=0; i<num; i++)
|
||||
tk::dnn::writeBUF(buf, mask[i]);
|
||||
for(int i=0; i<3*2*num; 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 classes, num;
|
||||
std::vector<std::string> classesNames;
|
||||
|
||||
dnnType *mask;
|
||||
dnnType *bias;
|
||||
|
||||
int entry_index(int batch, int location, int entry, int batchSize) {
|
||||
int n = location / (w*h);
|
||||
int loc = location % (w*h);
|
||||
return batch*c*h*w*batchSize + n*w*h*(4+classes+1) + entry*w*h + loc;
|
||||
}
|
||||
|
||||
};
|
||||
@@ -3,14 +3,6 @@
|
||||
*/
|
||||
#include "Network.h"
|
||||
#include "Layer.h"
|
||||
#include "NetworkRT.h"
|
||||
|
||||
namespace tkDNN {
|
||||
|
||||
/**
|
||||
Return the tkDNN version
|
||||
*/
|
||||
int getVersion() {
|
||||
|
||||
return 100;
|
||||
}
|
||||
}
|
||||
#define TKDNN_VERSION 400
|
||||
@@ -0,0 +1,116 @@
|
||||
#ifndef UTILS_H
|
||||
#define UTILS_H
|
||||
|
||||
#include <iostream>
|
||||
#include <sstream>
|
||||
#include <fstream>
|
||||
#include <iomanip>
|
||||
#include <stdlib.h>
|
||||
|
||||
#include "cuda.h"
|
||||
#include "cuda_runtime_api.h"
|
||||
#include <cublas_v2.h>
|
||||
#include <cudnn.h>
|
||||
|
||||
#define dnnType float
|
||||
|
||||
// Colored output
|
||||
#define COL_END "\033[0m"
|
||||
|
||||
#define COL_RED "\033[31m"
|
||||
#define COL_GREEN "\033[32m"
|
||||
#define COL_ORANGE "\033[33m"
|
||||
#define COL_BLUE "\033[34m"
|
||||
#define COL_PURPLE "\033[35m"
|
||||
#define COL_CYAN "\033[36m"
|
||||
|
||||
#define COL_REDB "\033[1;31m"
|
||||
#define COL_GREENB "\033[1;32m"
|
||||
#define COL_ORANGEB "\033[1;33m"
|
||||
#define COL_BLUEB "\033[1;34m"
|
||||
#define COL_PURPLEB "\033[1;35m"
|
||||
#define COL_CYANB "\033[1;36m"
|
||||
|
||||
// Simple Timer
|
||||
#define TIMER_START \
|
||||
timespec start, end; \
|
||||
clock_gettime(CLOCK_MONOTONIC, &start);
|
||||
|
||||
#define TIMER_STOP_C(col) \
|
||||
clock_gettime(CLOCK_MONOTONIC, &end); \
|
||||
double t_ns = ((double)(end.tv_sec - start.tv_sec) * 1.0e9 + \
|
||||
(double)(end.tv_nsec - start.tv_nsec)) / \
|
||||
1.0e6; \
|
||||
std::cout << col << "Time:" << std::setw(16) << t_ns << " ms\n" \
|
||||
<< COL_END;
|
||||
|
||||
#define TIMER_STOP TIMER_STOP_C(COL_CYANB)
|
||||
|
||||
/********************************************************
|
||||
* Prints the error message, and exits
|
||||
* ******************************************************/
|
||||
#define EXIT_WAIVED 0
|
||||
|
||||
#define FatalError(s) \
|
||||
{ \
|
||||
std::stringstream _where, _message; \
|
||||
_where << __FILE__ << ':' << __LINE__; \
|
||||
_message << std::string(s) + "\n" \
|
||||
<< __FILE__ << ':' << __LINE__; \
|
||||
std::cerr << _message.str() << "\nAborting...\n"; \
|
||||
cudaDeviceReset(); \
|
||||
exit(EXIT_FAILURE); \
|
||||
}
|
||||
|
||||
#define checkCUDNN(status) \
|
||||
{ \
|
||||
std::stringstream _error; \
|
||||
if (status != CUDNN_STATUS_SUCCESS) \
|
||||
{ \
|
||||
_error << "CUDNN failure: " << cudnnGetErrorString(status); \
|
||||
FatalError(_error.str()); \
|
||||
} \
|
||||
}
|
||||
|
||||
#define checkCuda(status) \
|
||||
{ \
|
||||
std::stringstream _error; \
|
||||
if (status != 0) \
|
||||
{ \
|
||||
_error << "Cuda failure: " << cudaGetErrorString(status); \
|
||||
FatalError(_error.str()); \
|
||||
} \
|
||||
}
|
||||
|
||||
#define checkERROR(status) \
|
||||
{ \
|
||||
std::stringstream _error; \
|
||||
if (status != 0) \
|
||||
{ \
|
||||
_error << "Generic failure: " << status; \
|
||||
FatalError(_error.str()); \
|
||||
} \
|
||||
}
|
||||
|
||||
#define checkNULL(ptr) \
|
||||
{ \
|
||||
std::stringstream _error; \
|
||||
if (ptr == nullptr) \
|
||||
{ \
|
||||
_error << "Null pointer"; \
|
||||
FatalError(_error.str()); \
|
||||
} \
|
||||
}
|
||||
|
||||
void printCenteredTitle(const char *title, char fill, int dim);
|
||||
bool fileExist(const char *fname);
|
||||
void readBinaryFile(std::string fname, int size, dnnType** data_h, dnnType** data_d, int seek = 0);
|
||||
int checkResult(int size, dnnType *data_d, dnnType *correct_d, bool device = true, int limit = 10);
|
||||
void printDeviceVector(int size, dnnType* vec_d, bool device = true);
|
||||
void resize(int size, dnnType **data);
|
||||
|
||||
void matrixTranspose(cublasHandle_t handle, dnnType *srcData, dnnType *dstData, int rows, int cols);
|
||||
|
||||
void matrixMulAdd(cublasHandle_t handle, dnnType *srcData, dnnType *dstData,
|
||||
dnnType *add_vector, int dim, dnnType mul);
|
||||
#endif //UTILS_H
|
||||