diff --git a/CMakeLists.txt b/CMakeLists.txt index f0d2ec1..88362f7 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -103,6 +103,9 @@ target_link_libraries(test_yolo3_flir tkDNN) add_executable(test_mobilenetv2ssd tests/mobilenetv2ssd/mobilenetv2ssd.cpp) target_link_libraries(test_mobilenetv2ssd tkDNN) +add_executable(test_mobilenetv2ssd512 tests/mobilenetv2ssd512/mobilenetv2ssd512.cpp) +target_link_libraries(test_mobilenetv2ssd512 tkDNN) + add_executable(test_resnet101 tests/resnet101/resnet101.cpp) target_link_libraries(test_resnet101 tkDNN) diff --git a/demo/demo/demo.cpp b/demo/demo/demo.cpp index a5a9d1a..90ff800 100644 --- a/demo/demo/demo.cpp +++ b/demo/demo/demo.cpp @@ -44,7 +44,7 @@ int main(int argc, char *argv[]) { cnet.init(net); break; case 'm': - mbnet.init(net); + mbnet.init(net, 512, 81); break; default: FatalError("Network type not allowed (3rd parameter)\n"); diff --git a/include/tkDNN/MobilenetDetection.h b/include/tkDNN/MobilenetDetection.h index 840a2cc..f7b45a4 100644 --- a/include/tkDNN/MobilenetDetection.h +++ b/include/tkDNN/MobilenetDetection.h @@ -15,31 +15,31 @@ struct SSDSpec { - int feature_size = 0; + int featureSize = 0; int shrinkage = 0; - int box_width = 0; - int box_height = 0; + int boxWidth = 0; + int boxHeight = 0; int ratio1 = 0; int ratio2 = 0; SSDSpec() {} - SSDSpec(int feature_size, int shrinkage, int box_width, int box_height, int ratio1, int ratio2) : feature_size(feature_size), shrinkage(shrinkage), box_width(box_width), box_height(box_height), + SSDSpec(int feature_size, int shrinkage, int box_width, int box_height, int ratio1, int ratio2) : featureSize(feature_size), shrinkage(shrinkage), boxWidth(box_width), boxHeight(box_height), ratio1(ratio1), ratio2(ratio2) {} void setAll(int feature_size, int shrinkage, int box_width, int box_height, int ratio1, int ratio2) { - this->feature_size = feature_size; + this->featureSize = feature_size; this->shrinkage = shrinkage; - this->box_width = box_width; - this->box_height = box_height; + this->boxWidth = box_width; + this->boxHeight = box_height; this->ratio1 = ratio1; this->ratio2 = ratio2; } void print() { - std::cout << "fsize: " << feature_size << "\tshrinkage: " << shrinkage << "\t box W:" << box_width << "\tbox H: " << box_height << "\t x ratio:" << ratio1 << "\t y ratio:" << ratio2 << std::endl; + std::cout << "fsize: " << featureSize << "\tshrinkage: " << shrinkage << "\t box W:" << boxWidth << "\tbox H: " << boxHeight << "\t x ratio:" << ratio1 << "\t y ratio:" << ratio2 << std::endl; } }; @@ -54,14 +54,12 @@ class MobilenetDetection private: tk::dnn::NetworkRT *netRT = nullptr; - int classes = 21; - float iou_threshold = 0.45; - float center_variance = 0.1; - float size_variance = 0.2; - float conf_thresh = 0.4; - int input_h = 300; - int input_w = 300; - int image_size = 300; + int classes; + float iouThreshold = 0.45; + float centerVariance = 0.1; + float sizeVariance = 0.2; + float confThresh = 0.4; + int imageSize; float *priors = nullptr; int n_priors = 0; @@ -91,7 +89,7 @@ private: float get_color2(int c, int x, int max); cv::Scalar colors[256]; - std::vector voc_class_name; + std::vector classesNames; public: // keep track of inference times (ms) @@ -101,7 +99,7 @@ public: MobilenetDetection() {} ~MobilenetDetection() {} - void init(std::string tensor_path); + void init(std::string tensor_path, int input_size, int n_classes); cv::Mat draw(); void update(cv::Mat &img); }; diff --git a/src/MobilenetDetection.cpp b/src/MobilenetDetection.cpp index 012ed9d..d68f346 100644 --- a/src/MobilenetDetection.cpp +++ b/src/MobilenetDetection.cpp @@ -15,7 +15,7 @@ void MobilenetDetection::generate_ssd_priors(const SSDSpec *specs, const int n_s n_priors = 0; for (int i = 0; i < n_specs; i++) { - n_priors += specs[i].feature_size * specs[i].feature_size * 6; + n_priors += specs[i].featureSize * specs[i].featureSize * 6; } // std::cout<<"n priors: "< specs[i].box_width ? specs[i].box_width : specs[i].box_height; - max = specs[i].box_height < specs[i].box_width ? specs[i].box_width : specs[i].box_height; - for (int j = 0; j < specs[i].feature_size; j++) + scale = (float)imageSize / (float)specs[i].shrinkage; + min = specs[i].boxHeight > specs[i].boxWidth ? specs[i].boxWidth : specs[i].boxHeight; + max = specs[i].boxHeight < specs[i].boxWidth ? specs[i].boxWidth : specs[i].boxHeight; + for (int j = 0; j < specs[i].featureSize; j++) { - for (int k = 0; k < specs[i].feature_size; k++) + for (int k = 0; k < specs[i].featureSize; k++) { //small sized square box size = min; x_center = (k + 0.5f) / scale; y_center = (j + 0.5f) / scale; - h = w = (float)size / (float)image_size; + h = w = (float)size / (float)imageSize; priors[i_prio * N_COORDS + 0] = x_center; priors[i_prio * N_COORDS + 1] = y_center; @@ -49,7 +49,7 @@ void MobilenetDetection::generate_ssd_priors(const SSDSpec *specs, const int n_s //big sized square box size = sqrt(max * min); - h = w = (float)size / (float)image_size; + h = w = (float)size / (float)imageSize; priors[i_prio * N_COORDS + 0] = x_center; priors[i_prio * N_COORDS + 1] = y_center; @@ -59,7 +59,7 @@ void MobilenetDetection::generate_ssd_priors(const SSDSpec *specs, const int n_s //change h/w ratio of the small sized box size = min; - h = w = size / (float)image_size; + h = w = size / (float)imageSize; ratio = sqrt(specs[i].ratio1); priors[i_prio * N_COORDS + 0] = x_center; priors[i_prio * N_COORDS + 1] = y_center; @@ -103,15 +103,15 @@ void MobilenetDetection::generate_ssd_priors(const SSDSpec *specs, const int n_s } } -void MobilenetDetection::convert_locatios_to_boxes_and_center(float *priors, const int n_priors, float *locations, const float center_variance, const float size_variance) +void MobilenetDetection::convert_locatios_to_boxes_and_center(float *priors, const int n_priors, float *locations, const float centerVariance, const float sizeVariance) { float cur_x, cur_y; for (int i = 0; i < n_priors; i++) { - locations[i * N_COORDS + 0] = locations[i * N_COORDS + 0] * center_variance * priors[i * N_COORDS + 2] + priors[i * N_COORDS + 0]; - locations[i * N_COORDS + 1] = locations[i * N_COORDS + 1] * center_variance * priors[i * N_COORDS + 3] + priors[i * N_COORDS + 1]; - locations[i * N_COORDS + 2] = exp(locations[i * N_COORDS + 2] * size_variance) * priors[i * N_COORDS + 2]; - locations[i * N_COORDS + 3] = exp(locations[i * N_COORDS + 3] * size_variance) * priors[i * N_COORDS + 3]; + locations[i * N_COORDS + 0] = locations[i * N_COORDS + 0] * centerVariance * priors[i * N_COORDS + 2] + priors[i * N_COORDS + 0]; + locations[i * N_COORDS + 1] = locations[i * N_COORDS + 1] * centerVariance * priors[i * N_COORDS + 3] + priors[i * N_COORDS + 1]; + locations[i * N_COORDS + 2] = exp(locations[i * N_COORDS + 2] * sizeVariance) * priors[i * N_COORDS + 2]; + locations[i * N_COORDS + 3] = exp(locations[i * N_COORDS + 3] * sizeVariance) * priors[i * N_COORDS + 3]; cur_x = locations[i * N_COORDS + 0]; cur_y = locations[i * N_COORDS + 1]; @@ -221,16 +221,38 @@ float MobilenetDetection::get_color2(int c, int x, int max) return r; } -void MobilenetDetection::init(std::string tensor_path) +void MobilenetDetection::init(std::string tensor_path, int input_size, int n_classes) { + this->imageSize = input_size; + this->classes = n_classes; + const int n_SSDSpec = 6; SSDSpec specs[6]; - specs[0].setAll(19, 16, 60, 105, 2, 3); - specs[1].setAll(10, 32, 105, 150, 2, 3); - specs[2].setAll(5, 64, 150, 195, 2, 3); - specs[3].setAll(3, 100, 195, 240, 2, 3); - specs[4].setAll(2, 150, 240, 285, 2, 3); - specs[5].setAll(1, 300, 285, 330, 2, 3); + + + if(input_size == 300) + { + specs[0].setAll(19, 16, 60, 105, 2, 3); + specs[1].setAll(10, 32, 105, 150, 2, 3); + specs[2].setAll(5, 64, 150, 195, 2, 3); + specs[3].setAll(3, 100, 195, 240, 2, 3); + specs[4].setAll(2, 150, 240, 285, 2, 3); + specs[5].setAll(1, 300, 285, 330, 2, 3); + } + else if(input_size == 512) + { + specs[0].setAll(32, 16, 60, 105, 2, 3); + specs[1].setAll(16, 32, 105, 150, 2, 3); + specs[2].setAll(8, 64, 150, 195, 2, 3); + specs[3].setAll(4, 100, 195, 240, 2, 3); + specs[4].setAll(2, 150, 240, 285, 2, 3); + specs[5].setAll(1, 300, 285, 330, 2, 3); + } + else + { + FatalError("Input size for mobilenet not supported"); + } + generate_ssd_priors(specs, n_SSDSpec); @@ -242,7 +264,7 @@ void MobilenetDetection::init(std::string tensor_path) locations_h = (float *)malloc(N_COORDS * n_priors * sizeof(float)); confidences_h = (float *)malloc(n_priors * classes * sizeof(float)); - dim = tk::dnn::dataDim_t(1, 3, input_w, input_h, 1); + dim = tk::dnn::dataDim_t(1, 3, imageSize, imageSize, 1); for (int c = 0; c < classes; c++) { @@ -253,11 +275,39 @@ void MobilenetDetection::init(std::string tensor_path) colors[c] = cv::Scalar(int(255.0 * b), int(255.0 * g), int(255.0 * r)); } - const char *voc_class_name_[] = { + if(classes == 21) + { + const char *classes_names_[] = { "BACKGROUND", "aeroplane", "bicycle", "bird", "boat", "bottle", "bus", "car", "cat", "chair", "cow", "diningtable", "dog", "horse", "motorbike", "person", "pottedplant", "sheep", "sofa", "train", "tvmonitor"}; - voc_class_name = std::vector(voc_class_name_, std::end(voc_class_name_)); + classesNames = std::vector(classes_names_, std::end(classes_names_)); + + } + else if (classes == 81) + { + const char *classes_names_[] = { + "BACKGROUND", "person" , "bicycle" , "car" , "motorbike" , "aeroplane" , "bus" , + "train" , "truck" , "boat" , "traffic light" , "fire hydrant" , "stop sign" , + "parking meter" , "bench" , "bird" , "cat" , "dog" , "horse" , "sheep" , "cow" , + "elephant" , "bear" , "zebra" , "giraffe" , "backpack" , "umbrella" , "handbag" , + "tie" , "suitcase" , "frisbee" , "skis" , "snowboard" , "sports ball" , "kite" , + "baseball bat" , "baseball glove" , "skateboard" , "surfboard" , "tennis racket" , + "bottle" , "wine glass" , "cup" , "fork" , "knife" , "spoon" , "bowl" , "banana" , + "apple" , "sandwich" , "orange" , "broccoli" , "carrot" , "hot dog" , "pizza" , + "donut" , "cake" , "chair" , "sofa" , "pottedplant" , "bed" , "diningtable" , + "toilet" , "tvmonitor" , "laptop" , "mouse" , "remote" , "keyboard" , + "cell phone" , "microwave" , "oven" , "toaster" , "sink" , "refrigerator" , + "book" , "clock" , "vase" , "scissors" , "teddy bear" , "hair drier" , "toothbrush"}; + classesNames = std::vector(classes_names_, std::end(classes_names_)); + + } + else + { + FatalError("Number of classes not supported for mobilenet"); + } + + } cv::Mat MobilenetDetection::draw() @@ -266,7 +316,7 @@ cv::Mat MobilenetDetection::draw() for (size_t i = 0; i < detected.size(); i++) { b = detected[i]; - std::string det_class = voc_class_name[b.cl]; + std::string det_class = classesNames[b.cl]; cv::rectangle(origImg, cv::Point(b.x, b.y), cv::Point(b.w, b.h), colors[b.cl], 2); // draw label cv::Size textSize = getTextSize(det_class, cv::FONT_HERSHEY_SIMPLEX, fontScale, thickness, &baseline); @@ -319,8 +369,8 @@ void MobilenetDetection::update(cv::Mat &img) checkCuda(cudaMemcpy(locations_h, loc, N_COORDS * n_priors * sizeof(float), cudaMemcpyDeviceToHost)); //postprocess - convert_locatios_to_boxes_and_center(priors, n_priors, locations_h, center_variance, size_variance); - detected = postprocess(locations_h, confidences_h, n_priors, conf_thresh, classes, iou_threshold, sz.width, sz.height); + convert_locatios_to_boxes_and_center(priors, n_priors, locations_h, centerVariance, sizeVariance); + detected = postprocess(locations_h, confidences_h, n_priors, confThresh, classes, iouThreshold, sz.width, sz.height); TIMER_STOP stats.push_back(t_ns); diff --git a/tests/mobilenetv2ssd512/mobilenetv2ssd512.cpp b/tests/mobilenetv2ssd512/mobilenetv2ssd512.cpp new file mode 100644 index 0000000..eb2887e --- /dev/null +++ b/tests/mobilenetv2ssd512/mobilenetv2ssd512.cpp @@ -0,0 +1,540 @@ +#include +#include "tkdnn.h" + + +const char *output_bin1 = "../tests/mobilenetv2ssd512/debug/classification_headers-5.bin"; +const char *output_bin2 = "../tests/mobilenetv2ssd512/debug/regression_headers-5.bin"; +const char *input_bin = "../tests/mobilenetv2ssd512/debug/input.bin"; + +const char *conv0_bin = "../tests/mobilenetv2ssd512/layers/base_net-0-0.bin"; +const char *inverted_residual1[] = { + "../tests/mobilenetv2ssd512/layers/base_net-1-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-1-conv-3.bin"}; +const char *inverted_residual2[] = { + "../tests/mobilenetv2ssd512/layers/base_net-2-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-2-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-2-conv-6.bin"}; +const char *inverted_residual3[] = { + "../tests/mobilenetv2ssd512/layers/base_net-3-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-3-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-3-conv-6.bin"}; +const char *inverted_residual4[] = { + "../tests/mobilenetv2ssd512/layers/base_net-4-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-4-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-4-conv-6.bin"}; +const char *inverted_residual5[] = { + "../tests/mobilenetv2ssd512/layers/base_net-5-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-5-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-5-conv-6.bin"}; +const char *inverted_residual6[] = { + "../tests/mobilenetv2ssd512/layers/base_net-6-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-6-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-6-conv-6.bin"}; +const char *inverted_residual7[] = { + "../tests/mobilenetv2ssd512/layers/base_net-7-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-7-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-7-conv-6.bin"}; +const char *inverted_residual8[] = { + "../tests/mobilenetv2ssd512/layers/base_net-8-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-8-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-8-conv-6.bin"}; +const char *inverted_residual9[] = { + "../tests/mobilenetv2ssd512/layers/base_net-9-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-9-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-9-conv-6.bin"}; +const char *inverted_residual10[] = { + "../tests/mobilenetv2ssd512/layers/base_net-10-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-10-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-10-conv-6.bin"}; +const char *inverted_residual11[] = { + "../tests/mobilenetv2ssd512/layers/base_net-11-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-11-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-11-conv-6.bin"}; +const char *inverted_residual12[] = { + "../tests/mobilenetv2ssd512/layers/base_net-12-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-12-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-12-conv-6.bin"}; +const char *inverted_residual13[] = { + "../tests/mobilenetv2ssd512/layers/base_net-13-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-13-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-13-conv-6.bin"}; +const char *inverted_residual14[] = { + "../tests/mobilenetv2ssd512/layers/base_net-14-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-14-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-14-conv-6.bin"}; +const char *inverted_residual15[] = { + "../tests/mobilenetv2ssd512/layers/base_net-15-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-15-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-15-conv-6.bin"}; +const char *inverted_residual16[] = { + "../tests/mobilenetv2ssd512/layers/base_net-16-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-16-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-16-conv-6.bin"}; +const char *inverted_residual17[] = { + "../tests/mobilenetv2ssd512/layers/base_net-17-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/base_net-17-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/base_net-17-conv-6.bin"}; + +const char *conv18 = "../tests/mobilenetv2ssd512/layers/base_net-18-0.bin"; + +const char *extras0[] = { + "../tests/mobilenetv2ssd512/layers/extras-0-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/extras-0-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/extras-0-conv-6.bin"}; +const char *extras1[] = { + "../tests/mobilenetv2ssd512/layers/extras-1-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/extras-1-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/extras-1-conv-6.bin"}; +const char *extras2[] = { + "../tests/mobilenetv2ssd512/layers/extras-2-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/extras-2-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/extras-2-conv-6.bin"}; +const char *extras3[] = { + "../tests/mobilenetv2ssd512/layers/extras-3-conv-0.bin", + "../tests/mobilenetv2ssd512/layers/extras-3-conv-3.bin", + "../tests/mobilenetv2ssd512/layers/extras-3-conv-6.bin"}; + +const char *classification_header0[] = { + "../tests/mobilenetv2ssd512/layers/classification_headers-0-0.bin", + "../tests/mobilenetv2ssd512/layers/classification_headers-0-3.bin"}; +const char *classification_header1[] = { + "../tests/mobilenetv2ssd512/layers/classification_headers-1-0.bin", + "../tests/mobilenetv2ssd512/layers/classification_headers-1-3.bin"}; +const char *classification_header2[] = { + "../tests/mobilenetv2ssd512/layers/classification_headers-2-0.bin", + "../tests/mobilenetv2ssd512/layers/classification_headers-2-3.bin"}; +const char *classification_header3[] = { + "../tests/mobilenetv2ssd512/layers/classification_headers-3-0.bin", + "../tests/mobilenetv2ssd512/layers/classification_headers-3-3.bin"}; +const char *classification_header4[] = { + "../tests/mobilenetv2ssd512/layers/classification_headers-4-0.bin", + "../tests/mobilenetv2ssd512/layers/classification_headers-4-3.bin"}; + +const char *classification_header5 = "../tests/mobilenetv2ssd512/layers/classification_headers-5.bin"; + +const char *regression_header0[] = { + "../tests/mobilenetv2ssd512/layers/regression_headers-0-0.bin", + "../tests/mobilenetv2ssd512/layers/regression_headers-0-3.bin"}; +const char *regression_header1[] = { + "../tests/mobilenetv2ssd512/layers/regression_headers-1-0.bin", + "../tests/mobilenetv2ssd512/layers/regression_headers-1-3.bin"}; +const char *regression_header2[] = { + "../tests/mobilenetv2ssd512/layers/regression_headers-2-0.bin", + "../tests/mobilenetv2ssd512/layers/regression_headers-2-3.bin"}; +const char *regression_header3[] = { + "../tests/mobilenetv2ssd512/layers/regression_headers-3-0.bin", + "../tests/mobilenetv2ssd512/layers/regression_headers-3-3.bin"}; +const char *regression_header4[] = { + "../tests/mobilenetv2ssd512/layers/regression_headers-4-0.bin", + "../tests/mobilenetv2ssd512/layers/regression_headers-4-3.bin"}; + +const char *regression_header5 = "../tests/mobilenetv2ssd512/layers/regression_headers-5.bin"; + + +int main() +{ + + // downloadWeightsifDoNotExist(input_bin, "./tests/mobilenetv2ssd512"); + + int classes = 81; + + // Network layout + tk::dnn::dataDim_t dim(1, 3, 512, 512, 1); + tk::dnn::Network net(dim); + + tk::dnn::Conv2d conv1(&net, 32, 3, 3, 2, 2, 1, 1, conv0_bin, true); + tk::dnn::Activation relu3(&net, CUDNN_ACTIVATION_RELU); + + //Inverted Residual 1 + + tk::dnn::Conv2d conv2(&net, 32, 3, 3, 1, 1, 1, 1, inverted_residual1[0], true, false, false, 32); + tk::dnn::Activation relu5(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d conv3(&net, 16, 1, 1, 1, 1, 0, 0, inverted_residual1[1], true); + + //Inverted Residual 2 + tk::dnn::Conv2d ir_2_conv1(&net, 96, 1, 1, 1, 1, 0, 0, inverted_residual2[0], true); + tk::dnn::Activation relu_2_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_2_conv2(&net, 96, 3, 3, 2, 2, 1, 1, inverted_residual2[1], true, false, false, 96); + tk::dnn::Activation relu_2_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_2_conv3(&net, 24, 1, 1, 1, 1, 0, 0, inverted_residual2[2], true); + + //Inverted Residual 3 + tk::dnn::Layer *last = &ir_2_conv3; + tk::dnn::Conv2d ir_3_conv1(&net, 144, 1, 1, 1, 1, 0, 0, inverted_residual3[0], true); + tk::dnn::Activation relu_3_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_3_conv2(&net, 144, 3, 3, 1, 1, 1, 1, inverted_residual3[1], true, false, false, 144); + tk::dnn::Activation relu_3_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_3_conv3(&net, 24, 1, 1, 1, 1, 0, 0, inverted_residual3[2], true); + + tk::dnn::Shortcut s3_0(&net, last); + // //Inverted Residual 4 + tk::dnn::Conv2d ir_4_conv1(&net, 144, 1, 1, 1, 1, 0, 0, inverted_residual4[0], true); + tk::dnn::Activation relu_4_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_4_conv2(&net, 144, 3, 3, 2, 2, 1, 1, inverted_residual4[1], true, false, false, 144); + tk::dnn::Activation relu_4_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_4_conv3(&net, 32, 1, 1, 1, 1, 0, 0, inverted_residual4[2], true); + + // // //Inverted Residual 5 + last = &ir_4_conv3; + tk::dnn::Conv2d ir_5_conv1(&net, 192, 1, 1, 1, 1, 0, 0, inverted_residual5[0], true); + tk::dnn::Activation relu_5_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_5_conv2(&net, 192, 3, 3, 1, 1, 1, 1, inverted_residual5[1], true, false, false, 192); + tk::dnn::Activation relu_5_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_5_conv3(&net, 32, 1, 1, 1, 1, 0, 0, inverted_residual5[2], true); + + tk::dnn::Shortcut s5_0(&net, last); + // // // //Inverted Residual 6 + last = &s5_0; + tk::dnn::Conv2d ir_6_conv1(&net, 192, 1, 1, 1, 1, 0, 0, inverted_residual6[0], true); + tk::dnn::Activation relu_6_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_6_conv2(&net, 192, 3, 3, 1, 1, 1, 1, inverted_residual6[1], true, false, false, 192); + tk::dnn::Activation relu_6_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_6_conv3(&net, 32, 1, 1, 1, 1, 0, 0, inverted_residual6[2], true); + + tk::dnn::Shortcut s6_0(&net, last); + //Inverted Residual 7 + tk::dnn::Conv2d ir_7_conv1(&net, 192, 1, 1, 1, 1, 0, 0, inverted_residual7[0], true); + tk::dnn::Activation relu_7_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_7_conv2(&net, 192, 3, 3, 2, 2, 1, 1, inverted_residual7[1], true, false, false, 192); + tk::dnn::Activation relu_7_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_7_conv3(&net, 64, 1, 1, 1, 1, 0, 0, inverted_residual7[2], true); + + // //Inverted Residual 8 + last = &ir_7_conv3; + tk::dnn::Conv2d ir_8_conv1(&net, 384, 1, 1, 1, 1, 0, 0, inverted_residual8[0], true); + tk::dnn::Activation relu_8_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_8_conv2(&net, 384, 3, 3, 1, 1, 1, 1, inverted_residual8[1], true, false, false, 384); + tk::dnn::Activation relu_8_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_8_conv3(&net, 64, 1, 1, 1, 1, 0, 0, inverted_residual8[2], true); + + tk::dnn::Shortcut s8_0(&net, last); + //Inverted Residual 9 + last = &s8_0; + tk::dnn::Conv2d ir_9_conv1(&net, 384, 1, 1, 1, 1, 0, 0, inverted_residual9[0], true); + tk::dnn::Activation relu_9_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_9_conv2(&net, 384, 3, 3, 1, 1, 1, 1, inverted_residual9[1], true, false, false, 384); + tk::dnn::Activation relu_9_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_9_conv3(&net, 64, 1, 1, 1, 1, 0, 0, inverted_residual9[2], true); + + tk::dnn::Shortcut s9_0(&net, last); + //Inverted Residual 10 + last = &s9_0; + tk::dnn::Conv2d ir_10_conv1(&net, 384, 1, 1, 1, 1, 0, 0, inverted_residual10[0], true); + tk::dnn::Activation relu_10_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_10_conv2(&net, 384, 3, 3, 1, 1, 1, 1, inverted_residual10[1], true, false, false, 384); + tk::dnn::Activation relu_10_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_10_conv3(&net, 64, 1, 1, 1, 1, 0, 0, inverted_residual10[2], true); + + tk::dnn::Shortcut s10_0(&net, last); + //Inverted Residual 11 + tk::dnn::Conv2d ir_11_conv1(&net, 384, 1, 1, 1, 1, 0, 0, inverted_residual11[0], true); + tk::dnn::Activation relu_11_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_11_conv2(&net, 384, 3, 3, 1, 1, 1, 1, inverted_residual11[1], true, false, false, 384); + tk::dnn::Activation relu_11_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_11_conv3(&net, 96, 1, 1, 1, 1, 0, 0, inverted_residual11[2], true); + + last = &ir_11_conv3; + //Inverted Residual 12 + tk::dnn::Conv2d ir_12_conv1(&net, 576, 1, 1, 1, 1, 0, 0, inverted_residual12[0], true); + tk::dnn::Activation relu_12_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_12_conv2(&net, 576, 3, 3, 1, 1, 1, 1, inverted_residual12[1], true, false, false, 576); + tk::dnn::Activation relu_12_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_12_conv3(&net, 96, 1, 1, 1, 1, 0, 0, inverted_residual12[2], true); + + tk::dnn::Shortcut s12_0(&net, last); + last = &s12_0; + //Inverted Residual 13 + tk::dnn::Conv2d ir_13_conv1(&net, 576, 1, 1, 1, 1, 0, 0, inverted_residual13[0], true); + tk::dnn::Activation relu_13_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_13_conv2(&net, 576, 3, 3, 1, 1, 1, 1, inverted_residual13[1], true, false, false, 576); + tk::dnn::Activation relu_13_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_13_conv3(&net, 96, 1, 1, 1, 1, 0, 0, inverted_residual13[2], true); + + tk::dnn::Shortcut s13_0(&net, last); + // //Inverted Residual 14 + tk::dnn::Conv2d ir_14_conv1(&net, 576, 1, 1, 1, 1, 0, 0, inverted_residual14[0], true); + tk::dnn::Activation relu_14_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_14_conv2(&net, 576, 3, 3, 2, 2, 1, 1, inverted_residual14[1], true, false, false, 576); + tk::dnn::Activation relu_14_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_14_conv3(&net, 160, 1, 1, 1, 1, 0, 0, inverted_residual14[2], true); + + // //Inverted Residual 15 + last = &ir_14_conv3; + tk::dnn::Conv2d ir_15_conv1(&net, 960, 1, 1, 1, 1, 0, 0, inverted_residual15[0], true); + tk::dnn::Activation relu_15_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_15_conv2(&net, 960, 3, 3, 1, 1, 1, 1, inverted_residual15[1], true, false, false, 960); + tk::dnn::Activation relu_15_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_15_conv3(&net, 160, 1, 1, 1, 1, 0, 0, inverted_residual15[2], true); + + tk::dnn::Shortcut s15_0(&net, last); + //Inverted Residual 16 + last = &s15_0; + tk::dnn::Conv2d ir_16_conv1(&net, 960, 1, 1, 1, 1, 0, 0, inverted_residual16[0], true); + tk::dnn::Activation relu_16_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_16_conv2(&net, 960, 3, 3, 1, 1, 1, 1, inverted_residual16[1], true, false, false, 960); + tk::dnn::Activation relu_16_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_16_conv3(&net, 160, 1, 1, 1, 1, 0, 0, inverted_residual16[2], true); + + tk::dnn::Shortcut s16_0(&net, last); + //Inverted Residual 17 + tk::dnn::Conv2d ir_17_conv1(&net, 960, 1, 1, 1, 1, 0, 0, inverted_residual17[0], true); + tk::dnn::Activation relu_17_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_17_conv2(&net, 960, 3, 3, 1, 1, 1, 1, inverted_residual17[1], true, false, false, 960); + tk::dnn::Activation relu_17_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d ir_17_conv3(&net, 320, 1, 1, 1, 1, 0, 0, inverted_residual17[2], true); + + //Conv 18 + tk::dnn::Conv2d ir_18_conv1(&net, 1280, 1, 1, 1, 1, 0, 0, conv18, true); + tk::dnn::Activation relu_18_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Layer *header_1[1] = {&relu_18_1}; + + // //extras Inverted Residual 0 + tk::dnn::Conv2d e_0_conv1(&net, 256, 1, 1, 1, 1, 0, 0, extras0[0], true); + tk::dnn::Activation e_relu_0_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_0_conv2(&net, 256, 3, 3, 2, 2, 1, 1, extras0[1], true, false, false, 256); + tk::dnn::Activation e_relu_0_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_0_conv3(&net, 512, 1, 1, 1, 1, 0, 0, extras0[2], true); + tk::dnn::Layer *header_2[1] = {&e_0_conv3}; + + // //extras Inverted Residual 1 + tk::dnn::Conv2d e_1_conv1(&net, 128, 1, 1, 1, 1, 0, 0, extras1[0], true); + tk::dnn::Activation e_relu_1_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_1_conv2(&net, 128, 3, 3, 2, 2, 1, 1, extras1[1], true, false, false, 128); + tk::dnn::Activation e_relu_1_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_1_conv3(&net, 256, 1, 1, 1, 1, 0, 0, extras1[2], true); + tk::dnn::Layer *header_3[1] = {&e_1_conv3}; + + //extras Inverted Residual 2 + tk::dnn::Conv2d e_2_conv1(&net, 128, 1, 1, 1, 1, 0, 0, extras2[0], true); + tk::dnn::Activation e_relu_2_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_2_conv2(&net, 128, 3, 3, 2, 2, 1, 1, extras2[1], true, false, false, 128); + tk::dnn::Activation e_relu_2_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_2_conv3(&net, 256, 1, 1, 1, 1, 0, 0, extras2[2], true); + tk::dnn::Layer *header_4[1] = {&e_2_conv3}; + + //extras Inverted Residual 3 + tk::dnn::Conv2d e_3_conv1(&net, 64, 1, 1, 1, 1, 0, 0, extras3[0], true); + tk::dnn::Activation e_relu_3_1(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_3_conv2(&net, 64, 3, 3, 2, 2, 1, 1, extras3[1], true, false, false, 64); + tk::dnn::Activation e_relu_3_2(&net, CUDNN_ACTIVATION_RELU); + tk::dnn::Conv2d e_3_conv3(&net, 64, 1, 1, 1, 1, 0, 0, extras3[2], true); + tk::dnn::Layer *header_5[1] = {&e_3_conv3}; + + // classification header 0 + tk::dnn::Layer *header_0[1] = {&relu_14_1}; + tk::dnn::Route rout_ch_0(&net, header_0, 1); + tk::dnn::Conv2d ch_0_conv1(&net, 576, 3, 3, 1, 1, 1, 1, classification_header0[0], true, false, false, 576, true); + tk::dnn::Activation ch_relu_0_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d ch_0_conv2(&net, 486, 1, 1, 1, 1, 0, 0, classification_header0[1], false); + tk::dnn::Layer *conf0[1] = {&ch_0_conv2}; + + // // classification header 1 + tk::dnn::Route rout_ch_1(&net, header_1, 1); + tk::dnn::Conv2d ch_1_conv1(&net, 1280, 3, 3, 1, 1, 1, 1, classification_header1[0], true, false, false, 1280, true); + tk::dnn::Activation ch_relu_1_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d ch_1_conv2(&net, 486, 1, 1, 1, 1, 0, 0, classification_header1[1], false); + tk::dnn::Layer *conf1[1] = {&ch_1_conv2}; + + // //classification header 2 + tk::dnn::Route rout_ch_2(&net, header_2, 1); + tk::dnn::Conv2d ch_2_conv1(&net, 512, 3, 3, 1, 1, 1, 1, classification_header2[0], true, false, false, 512, true); + tk::dnn::Activation ch_relu_2_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d ch_2_conv2(&net, 486, 1, 1, 1, 1, 0, 0, classification_header2[1], false); + tk::dnn::Layer *conf2[1] = {&ch_2_conv2}; + + // //classification header 3 + tk::dnn::Route rout_ch_3(&net, header_3, 1); + tk::dnn::Conv2d ch_3_conv1(&net, 256, 3, 3, 1, 1, 1, 1, classification_header3[0], true, false, false, 256, true); + tk::dnn::Activation ch_relu_3_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d ch_3_conv2(&net, 486, 1, 1, 1, 1, 0, 0, classification_header3[1], false); + tk::dnn::Layer *conf3[1] = {&ch_3_conv2}; + + // //classification header 4 + tk::dnn::Route rout_ch_4(&net, header_4, 1); + tk::dnn::Conv2d ch_4_conv1(&net, 256, 3, 3, 1, 1, 1, 1, classification_header4[0], true, false, false, 256, true); + tk::dnn::Activation ch_relu_4_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d ch_4_conv2(&net, 486, 1, 1, 1, 1, 0, 0, classification_header4[1], false); + tk::dnn::Layer *conf4[1] = {&ch_4_conv2}; + + // //classification header 5 + tk::dnn::Route rout_ch_5(&net, header_5, 1); + tk::dnn::Conv2d ch_5_conv(&net, 486, 1, 1, 1, 1, 0, 0, classification_header5, false, false, true); + tk::dnn::Layer *conf5[1] = {&ch_5_conv}; + + //regression header 0 + tk::dnn::Route rout_rh_0(&net, header_0, 1); + tk::dnn::Conv2d rh_0_conv1(&net, 576, 3, 3, 1, 1, 1, 1, regression_header0[0], true, false, false, 576, true); + tk::dnn::Activation rh_relu_0_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d rh_0_conv2(&net, 24, 1, 1, 1, 1, 0, 0, regression_header0[1], false); + tk::dnn::Layer *loc0[1] = {&rh_0_conv2}; + + // //regression header 1 + tk::dnn::Route rout_rh_1(&net, header_1, 1); + tk::dnn::Conv2d rh_1_conv1(&net, 1280, 3, 3, 1, 1, 1, 1, regression_header1[0], true, false, false, 1280, true); + tk::dnn::Activation rh_relu_1_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d rh_1_conv2(&net, 24, 1, 1, 1, 1, 0, 0, regression_header1[1], false); + tk::dnn::Layer *loc1[1] = {&rh_1_conv2}; + + //regression header 2 + tk::dnn::Route rout_rh_2(&net, header_2, 1); + tk::dnn::Conv2d rh_2_conv1(&net, 512, 3, 3, 1, 1, 1, 1, regression_header2[0], true, false, false, 512, true); + tk::dnn::Activation rh_relu_2_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d rh_2_conv2(&net, 24, 1, 1, 1, 1, 0, 0, regression_header2[1], false); + tk::dnn::Layer *loc2[1] = {&rh_2_conv2}; + + //regression header 3 + tk::dnn::Route rout_rh_3(&net, header_3, 1); + tk::dnn::Conv2d rh_3_conv1(&net, 256, 3, 3, 1, 1, 1, 1, regression_header3[0], true, false, false, 256, true); + tk::dnn::Activation rh_relu_3_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d rh_3_conv2(&net, 24, 1, 1, 1, 1, 0, 0, regression_header3[1], false); + tk::dnn::Layer *loc3[1] = {&rh_3_conv2}; + + //regression header 4 + + tk::dnn::Route rout_rh_4(&net, header_4, 1); + tk::dnn::Conv2d rh_4_conv1(&net, 256, 3, 3, 1, 1, 1, 1, regression_header4[0], true, false, false, 256, true); + tk::dnn::Activation rh_relu_4_1(&net, CUDNN_ACTIVATION_CLIPPED_RELU, 6); + tk::dnn::Conv2d rh_4_conv2(&net, 24, 1, 1, 1, 1, 0, 0, regression_header4[1], false); + tk::dnn::Layer *loc4[1] = {&rh_4_conv2}; + + //regression header 5 + tk::dnn::Route rout_rh_5(&net, header_5, 1); + tk::dnn::Conv2d rh_5_conv(&net, 24, 1, 1, 1, 1, 0, 0, regression_header5, false, false, true); + tk::dnn::Layer *loc5[1] = {&rh_5_conv}; + + last = &rh_5_conv; + + //flatten all confidence + tk::dnn::Route r_conf_0(&net, conf0, 1); + tk::dnn::Flatten fl_c_0(&net); + tk::dnn::Route r_conf_1(&net, conf1, 1); + tk::dnn::Flatten fl_c_1(&net); + tk::dnn::Route r_conf_2(&net, conf2, 1); + tk::dnn::Flatten fl_c_2(&net); + tk::dnn::Route r_conf_3(&net, conf3, 1); + tk::dnn::Flatten fl_c_3(&net); + tk::dnn::Route r_conf_4(&net, conf4, 1); + tk::dnn::Flatten fl_c_4(&net); + tk::dnn::Route r_conf_5(&net, conf5, 1); + tk::dnn::Flatten fl_c_5(&net); + + // //flatten all locations + tk::dnn::Route r_loc_0(&net, loc0, 1); + tk::dnn::Flatten fl_l_0(&net); + tk::dnn::Route r_loc_1(&net, loc1, 1); + tk::dnn::Flatten fl_l_1(&net); + tk::dnn::Route r_loc_2(&net, loc2, 1); + tk::dnn::Flatten fl_l_2(&net); + tk::dnn::Route r_loc_3(&net, loc3, 1); + tk::dnn::Flatten fl_l_3(&net); + tk::dnn::Route r_loc_4(&net, loc4, 1); + tk::dnn::Flatten fl_l_4(&net); + tk::dnn::Route r_loc_5(&net, loc5, 1); + tk::dnn::Flatten fl_l_5(&net); + + // //concat confidence + softmax + tk::dnn::Layer *confidences[6] = {&fl_c_0, &fl_c_1, &fl_c_2, &fl_c_3, &fl_c_4, &fl_c_5}; + tk::dnn::Route rout_conf(&net, confidences, 6); + tk::dnn::dataDim_t olddim_c = net.layers[net.num_layers - 1]->output_dim; + tk::dnn::dataDim_t dim_resh(1, olddim_c.c * olddim_c.h * olddim_c.w / classes, classes, 1, 1); + + tk::dnn::Reshape reshape_conf1(&net, dim_resh); + tk::dnn::Flatten fl_l_6(&net); + tk::dnn::dataDim_t newdim_c(1, classes, olddim_c.c * olddim_c.h * olddim_c.w / classes, 1, 1); + + tk::dnn::Reshape reshape_conf2(&net, newdim_c); + + tk::dnn::Softmax sm_1(&net, &newdim_c, true); + // tk::dnn::Flatten fl_l_7(&net); + // tk::dnn::Reshape reshape_conf3(&net,dim_resh, true); + tk::dnn::Layer *conf = &sm_1; + + //concat locations + tk::dnn::Layer *locations[6] = {&fl_l_0, &fl_l_1, &fl_l_2, &fl_l_3, &fl_l_4, &fl_l_5}; + tk::dnn::Route rout_loc(&net, locations, 6); + tk::dnn::dataDim_t olddim_l = net.layers[net.num_layers - 1]->output_dim; + tk::dnn::dataDim_t newdim_l(1, olddim_l.c * olddim_l.h * olddim_l.w / 4, 1, 4, 1); + tk::dnn::Reshape reshape_loc(&net, newdim_l, true); + tk::dnn::Layer *loc = &reshape_loc; + + // Load input + dnnType *data; + dnnType *input_h; + readBinaryFile(input_bin, dim.tot(), &input_h, &data); + //printDeviceVector(64, data, true); + + //print network model + net.print(); + + // convert network to tensorRT + tk::dnn::NetworkRT netRT(&net, "mobilenetv2ssd512.rt"); + + tk::dnn::dataDim_t dim1 = dim; //input dim + printCenteredTitle(" CUDNN inference ", '=', 30); + { + dim1.print(); + TIMER_START + net.infer(dim1, data); + TIMER_STOP + dim1.print(); + } + + dnnType *cudnn_out1 = conf5[0]->dstData; + tk::dnn::dataDim_t out_dim1 = conf5[0]->output_dim; + dnnType *cudnn_out2 = loc5[0]->dstData; + tk::dnn::dataDim_t out_dim2 = loc5[0]->output_dim; + + tk::dnn::dataDim_t dim2 = dim; + printCenteredTitle(" TENSORRT inference ", '=', 30); + { + dim2.print(); + TIMER_START + netRT.infer(dim2, data); + TIMER_STOP + dim2.print(); + } + + dnnType *rt_out1 = (dnnType *)netRT.buffersRT[1]; + dnnType *rt_out2 = (dnnType *)netRT.buffersRT[2]; + dnnType *rt_out3 = (dnnType *)netRT.buffersRT[3]; + dnnType *rt_out4 = (dnnType *)netRT.buffersRT[4]; + + printCenteredTitle(std::string(" RESNET CHECK RESULTS ").c_str(), '=', 30); + dnnType *out1, *out1_h; + int odim1 = out_dim1.tot(); + readBinaryFile(output_bin1, odim1, &out1_h, &out1); + + dnnType *out2, *out2_h; + int odim2 = out_dim2.tot(); + readBinaryFile(output_bin2, odim2, &out2_h, &out2); + std::cout << "CUDNN vs correct" << std::endl; + checkResult(odim1, cudnn_out1, out1); + checkResult(odim2, cudnn_out2, out2); + + std::cout << "TRT vs correct" << std::endl; + checkResult(odim1, rt_out1, out1); + checkResult(odim2, rt_out2, out2); + + std::cout << "CUDNN vs TRT " << std::endl; + checkResult(odim1, cudnn_out1, rt_out1); + checkResult(odim2, cudnn_out2, rt_out2); + + std::cout << "---------------------------------------------------" << std::endl; + std::cout << "Confidence CUDNN" << std::endl; + printDeviceVector(64, conf->dstData, true); + std::cout << "Locations CUDNN" << std::endl; + printDeviceVector(64, loc->dstData, true); + std::cout << "---------------------------------------------------" << std::endl; + + std::cout << "Confidence tensorRT" << std::endl; + printDeviceVector(64, rt_out3, true); + std::cout << "Locations tensorRT" << std::endl; + printDeviceVector(64, rt_out4, true); + std::cout << "---------------------------------------------------" << std::endl; + + std::cout << "CUDNN vs TRT " << std::endl; + checkResult(conf->output_dim.tot(), conf->dstData, rt_out3); + checkResult(loc->output_dim.tot(), loc->dstData, rt_out4); + + return 0; +}