RepVGG: Making VGG-style ConvNets Great Again

论文下载地址:https://arxiv.org/abs/2101.03697
官方源码(Pytorch实现):https://github.com/DingXiaoH/RepVGG

这篇论文对于我来说最大的用处是提出了结构的重重参数化:

在推理时将三个并行分支合并成单个分支,并保证输出输出不变。

结构重参数化主要分为两步,第一步主要是将Conv2d算子和BN算子融合以及将只有BN的分支转换成一个Conv2d算子,第二步将每个分支上的3x3卷积层融合成一个卷积层。

1、Conv2d和BN 这个已经是非常常见的,因为卷积核bn都是线性运算,所以可以进行合并。

这里假设输入的特征图(Input feature map)如下图所示,输入通道数为2,然后采用两个卷积核(图中只画了第一个卷积核对应参数)。

在这里插入图片描述

接着计算一下输出特征图(Output feature map)通道1上的第一个元素,即当卷积核1在输入特征图红色框区域卷积时得到的值(为了保证输入输出特征图高宽不变,所以对Input feature map进行了Padding)。其他位置的计算过程类似这里就不去演示了。

在这里插入图片描述

然后再将卷积层输出的特征图作为BN层的输入,这里同样计算一下输出特征图(Output feature map)通道1上的第一个元素,按照上述BN在推理时的计算公式即可得到如下图所示的计算结果。

在这里插入图片描述

代码

Conv2d+BN融合实验(Pytorch)
下面是参考作者提供的源码改的一个小实验,首先创建了一个module包含了卷积和BN模块,然后按照上述转换公式将卷积层的权重和BN的权重进行融合转换,接着载入到新建的卷积模块fused_conv中,最后随机创建一个Tensor(f1)将它分别输入到module以及fused_conv中,通过对比两者的输出可以发现它们的结果是一致的。

from collections import OrderedDict

import numpy as np
import torch
import torch.nn as nn


def main():
    torch.random.manual_seed(0)

    f1 = torch.randn(1, 2, 3, 3)

    module = nn.Sequential(OrderedDict(
        conv=nn.Conv2d(in_channels=2, out_channels=2, kernel_size=3, stride=1, padding=1, bias=False),
        bn=nn.BatchNorm2d(num_features=2)
    ))

    module.eval()

    with torch.no_grad():
        output1 = module(f1)
        print(output1)

    # fuse conv + bn
    kernel = module.conv.weight 
    running_mean = module.bn.running_mean
    running_var = module.bn.running_var
    gamma = module.bn.weight
    beta = module.bn.bias
    eps = module.bn.eps
    std = (running_var + eps).sqrt()
    t = (gamma / std).reshape(-1, 1, 1, 1)  # [ch] -> [ch, 1, 1, 1]
    kernel = kernel * t
    bias = beta - running_mean * gamma / std
    fused_conv = nn.Conv2d(in_channels=2, out_channels=2, kernel_size=3, stride=1, padding=1, bias=True)
    fused_conv.load_state_dict(OrderedDict(weight=kernel, bias=bias))

    with torch.no_grad():
        output2 = fused_conv(f1)
        print(output2)

    np.testing.assert_allclose(output1.numpy(), output2.numpy(), rtol=1e-03, atol=1e-05)
    print("convert module has been tested, and the result looks good!")


if __name__ == '__main__':
    main()

repVGG中大量运用conv+BN层,我们知道将层合并,减少层数能提升网络性能,下面的推理是conv带有bias的过程:

这其实就是一个卷积层,只不过权重考虑了BN的参数 我们令:

最终的融合结果即为:

相关融合代码如下图所示:

def _fuse_bn_tensor(self, branch):
        if branch is None:
            return 0, 0
        if isinstance(branch, nn.Sequential):
            kernel = branch.conv.weight
            running_mean = branch.bn.running_mean
            running_var = branch.bn.running_var
            gamma = branch.bn.weight
            beta = branch.bn.bias
            eps = branch.bn.eps
        else:
            ...
        std = (running_var + eps).sqrt()
        t = (gamma / std).reshape(-1, 1, 1, 1)
        return kernel * t, beta - running_mean * gamma / std

2、如何将不同分支合并:

作者这里首先将不同分支的卷积核都变成3*3:

2.1 将1×1卷积转换成3×3卷积
这个过程比较简单,如下图所示,以1×1卷积层中某一个卷积核为例,只需在原来权重周围补一圈零就行了,这样就变成了3×3的卷积层,注意为了保证输入输出特征图高宽不变,此时需要将padding设置成1(原来卷积核大小为1×1时padding为0)。最后按照上述2.1中讲的内容将卷积层和BN层进行融合即可。

在这里插入图片描述

2.2将BN转换成3×3卷积
对于只有BN的分支由于没有卷积层,所以我们可以先自己构建出一个卷积层来。如下图所示,构建了一个3×3的卷积层,该卷积层只做了恒等映射,即输入输出特征图不变。既然有了卷积层,那么又可以按照上述2.1中讲的内容将卷积层和BN层进行融合。

在这里插入图片描述

2.3 多分支融合
在上面的章节中,我们已经讲了怎么把每个分支融合转换成一个3×3的卷积层,接下来需要进一步将多分支转换成一个单路3×3卷积层。

在这里插入图片描述

合并的过程其实也很简单,直接将这三个卷积层的参数相加即可,具体推理过程就不讲了,如果不了解的可以自己动手算算。

总的来说,这篇论文的目标是Simple is Fast, Memory-economical, Flexible,提出了很多想法去实现上述目标,对于当前我的工作还是比较有启发的,尤其是最后对网络进行合并以及量化部分。下一步要好好学习下torch的量化QAT (torch.quantization.prepare_qat)

Verilog 阻塞赋值和非阻塞赋值

来源:《Verilog数字系统设计(夏宇闻)》

  • 在描述组合逻辑的always块中用阻塞赋值(=),则综合成组合逻辑的电路结构。
  • 在描述时序逻辑的always块中用非阻塞赋值(<=),则综合成时序逻辑的电路结构。

为什么一定要这样做呢?因为要使综合前仿真和综合后仿真一致的缘故。如果不按照上面两个要点来编写Verilog代码,也有可能综合出正确的逻辑,但前后仿真的结果就会不一致。

为了更好地理解上述要点,我们需要对Verilog 语言中的阻塞赋值和非阻塞赋值的功能和执行时间上的差别有深入的了解。为了解释问题方便下面定义两个缩写字:

RHS – 方程式右手方向的表达式或变量可分别缩写为:RHS表达式或RHS变量。

LHS – 方程式左手方向的表达式或变量可分别缩写为:LHS表达式或LHS变量。

IEEE Verilog标准定义了有些语句有确定的执行时间,有些语句没有确定的执行时间。若有两条或两条以上语句准备在同一时刻执行,但由于语句的排列次序不同(而这种排列次序的不同是IEEE Verilog标准所允许的), 却产生了不同的输出结果。这就是造成Verilog模块冒险和竞争现象的原因。为了避免产生竞争,理解阻塞和非阻塞赋值在执行时间上的差别是至关重要的。

阻塞赋值

阻塞赋值操作符用等号(即 = )表示。为什么称这种赋值为阻塞赋值呢?这是因为在赋值时先计算等号右手方向(RHS)部分的值,这时赋值语句不允许任何别的Verilog语句的干扰,直到现行的赋值完成时刻,即把RHS赋值给 LHS的时刻,它才允许别的赋值语句的执行。

一般可综合的阻塞赋值操作在RHS不能设定有延迟,(即使是零延迟也不允许)。从理论上讲,它与后面的赋值语句只有概念上的先后,而无实质上的延迟。若在RHS 加上延迟,则在延迟期间会阻止赋值语句的执行, 延迟后才执行赋值,这种赋值语句是不可综合的,在需要综合的模块设计中不可使用这种风格的代码。

阻塞赋值的执行可以认为是只有一个步骤的操作:

计算RHS并更新LHS,此时不能允许有来自任何其他Verilog语句的干扰。所谓阻塞的概念是指在同一个always块中,其后面的赋值语句从概念上(即使不设定延迟)是在前一句赋值语句结束后再开始赋值的。

如果在一个过程块中阻塞赋值的RHS变量正好是另一个过程块中阻塞赋值的LHS变量,这两个过程块又用同一个时钟沿触发,这时阻塞赋值操作会出现问题,即如果阻塞赋值的次序安排不好,就会出现竞争。若这两个阻塞赋值操作用同一个时钟沿触发,则执行的次序是无法确定的。下面的例子可以说明这个问题:

[例1]. 用阻塞赋值的反馈振荡器
    module fbosc1 (y1, y2, clk, rst);
      output y1, y2;
      input  clk, rst;
      reg    y1, y2;

      always @(posedge clk or posedge rst)
        if (rst) y1 = 0;  // reset
        else     y1 = y2;

      always @(posedge clk or posedge rst)
        if (rst) y2 = 1;  // preset
        else     y2 = y1;
    endmodule

按照IEEE Verilog 的标准,上例中两个always块是并行执行的,与前后次序无关。如果前一个always块的复位信号先到0时刻,则y1 和y2都会取1,而如果后一个always块的复位信号先到0时刻,则y1 和y2都会取0。这清楚地说明这个Verilog模块是不稳定的会产生冒险和竞争的情况。

非阻塞赋值

非阻塞赋值操作符用小于等于号 (即 <= )表示。为什么称这种赋值为非阻塞赋值?这是因为在赋值操作时刻开始时计算非阻塞赋值符的RHS表达式,赋值操作时刻结束时更新LHS。在计算非阻塞赋值的RHS表达式和更新LHS期间,其他的Verilog语句,包括其他的Verilog非阻塞赋值语句都能同时计算RHS表达式和更新LHS。非阻塞赋值允许其他的Verilog语句同时进行操作。非阻塞赋值的操作可以看作为两个步骤的过程:

  1. 在赋值时刻开始时,计算非阻塞赋值RHS表达式。
  2. 在赋值时刻结束时,更新非阻塞赋值LHS表达式。

非阻塞赋值操作只能用于对寄存器类型变量进行赋值,因此只能用在”initial”块和”always”块等过程块中。非阻塞赋值不允许用于连续赋值。下面的例子可以说明这个问题:

[例2]. 用非阻塞赋值的反馈振荡器
    module fbosc2 (y1, y2, clk, rst);
      output y1, y2;
      input  clk, rst;
      reg    y1, y2;

      always @(posedge clk or posedge rst)
        if (rst) y1 <= 0;  // reset
        else     y1 <= y2;

      always @(posedge clk or posedge rst)
        if (rst) y2 <= 1;  // preset
        else     y2 <= y1;
    endmodule

同样,按照IEEE Verilog 的标准,上例中两个always块是并行执行的,与前后次序无关。无论哪一个always块的复位信号先到, 两个always块中的非阻塞赋值都在赋值开始时刻计算RHS表达式,,而在结束时刻才更新LHS表达式。所以这两个always块在复位信号到来后,在always块结束时 y1为0而y2为1是确定的。从用户的角度看这两个非阻塞赋值正好是并行执行的。

Verilog模块编程要点:

下面我们还将对阻塞和非阻塞赋值做进一步解释并将举更多的例子来说明这个问题。在此之前,掌握可综合风格的Verilog模块编程的八个原则会有很大的帮助。在编写时牢记这八个要点可以为绝大多数的Verilog用户解决在综合后仿真中出现的90-100% 的冒险竞争问题。

  • 时序电路建模时,用非阻塞赋值。
  • 锁存器电路建模时,用非阻塞赋值。
  • 用always块建立组合逻辑模型时,用阻塞赋值。
  • 在同一个always块中建立时序和组合逻辑电路时,用非阻塞赋值。
  • 在同一个always块中不要既用非阻塞赋值又用阻塞赋值。
  • 不要在一个以上的always块中为同一个变量赋值。
  • 用$strobe系统任务来显示用非阻塞赋值的变量值
  • 在赋值时不要使用 #0 延迟

Verilog的新用户在彻底搞明白这两种赋值功能差别之前,一定要牢记这几条要点。照着要点来编写Verilog模块程序,就可省去很多麻烦。

yolov3 -tiny 网络实现和源码分析

摘自:https://blog.csdn.net/alangaixiaoxiao/article/details/105533746

相关推荐

网上的资源很多,也有很多博主的原理讲解。在这里推荐几个资源:
darknet官网:https://pjreddie.com/darknet/yolo/ linux系统
github上基于VS的工程:https://github.com/AlexeyAB/darknet
github上带有注释的工程:https://github.com/hgpvision/darknet

yolov3-tiny 原理

Yolo算法采用一个单独的CNN模型实现end-to-end的目标检测,首先将输入图片resize到448×448,然后送入CNN网络,最后处理网络预测结果得到检测的目标。
YOLO 的核心思想就是利用整张图作为网络的输入,直接在输出层回归 bounding box(边界框) 的位置及其所属的类别。将一幅图像分成 SxS 个网格(grid cell),如果某个 object 的中心落在这个网格中,则这个网格就负责预测这个 object。

每个 bounding box 要预测 (x, y, w, h) 和 confidence 共5个值,每个网格还要预测一个类别信息,记为 C 类。则 SxS个 网格,每个网格要预测 B 个 bounding box, 每个box中都有 C 个 classes对应的概率值。输出就是 S x S x B x(5+C) 的一个 tensor。

注意:class 信息是针对每个网格的,confidence 信息是针对每个 bounding box 的。

yolov3-tiny中,共有两个输出层(yolo层),分别为13×13和26×26,每个网格可以预测3个bounding box,共有80个分类数。所以最后的yolo层的尺寸为:13x13x255和26x26x255。
yolov3-tiny网络层结构如下:

可以看出,yolov3-tiny共有23层网络,其中包含五种不同的网络层:卷积层convolutional(13个),池化层maxpool(6个),路由层route(2个),上采样层upsample(1个),输出层yolo(2个)。Yolov3-tiny中,除了Yolo层之前的那个卷积层,每个卷积层之后都有BN层,且每个卷积层之后都有激活函数LEAKY(yolo层之前是linear)。

yolov3-tiny 源码分析

配置网络结构

yolov3-tiny前向传播主要在detector.c中的test_detector函数中完成:

/** 本函数是检测模型的一个前向推理测试函数.
* @param datacfg       数据集信息文件路径(也即cfg/*.data文件),文件中包含有关数据集的信息,比如cfg/coco.data
* @param cfgfile       网络配置文件路径(也即cfg/*.cfg文件),包含一个网络所有的结构参数,比如cfg/yolo.cfg
* @param weightfile    已经训练好的网络权重文件路径,比如darknet网站上下载的yolo.weights文件
* @param filename      待进行检测的图片路径(单张图片)
* @param thresh        阈值,类别检测概率大于该阈值才认为其检测结果有效
* @param hier_thresh
* @param outfile
* @param fullscreen
* @details 该函数为一个前向推理测试函数,不包括训练过程,因此如果要使用该函数,必须提前训练好网络,并加载训练好的网络参数文件,
*          这些文件可以在作者网站上根据作者的提示下载到。本函数由darknet.c中的主函数调用,严格来说,本文件不应纳入darknet网络结构文件夹中,
*          其只是一个测试文件,或者说是一个example,应该放入到example文件夹中(新版的darknet已经这样做了,可以在github上查看)。
*          本函数的流程为:.
*/
void test_detector(char *datacfg, char *cfgfile, char *weightfile, char *filename, float thresh,
    float hier_thresh, int dont_show, int ext_output, int save_labels, char *outfile, int letter_box, int benchmark_layers)
{
	// 从指定数据文件datacfg(.data文件)中读入数据信息(测试、训练数据信息)到options中
	// options是list类型数据,其中的node包含的void指针具体是kvp数据类型,具有键值和值(类似C++中的Map)
    list *options = read_data_cfg(datacfg);
	// 获取数据集的名称(包括路径),第二个参数"names"表明要从options中获取所用数据集的名称信息(如names = data/coco.names)
    char *name_list = option_find_str(options, "names", "data/names.list");
    int names_size = 0;
	// 从data/**.names中读取物体名称/标签信息
    char **names = get_labels_custom(name_list, &names_size); //get_labels(name_list);
	
    // 加载data/labels/文件夹中所有的字符标签图片
    image **alphabet = load_alphabet();

    network net = parse_network_cfg_custom(cfgfile, 1, 1); // set batch=1  配置各网络层参数,重要

在parser.c中的parse_network_cfg_custom函数中,根据yolov3-tiny.cfg文件对网络结构进行配置,明确各层网络的类型、输入输出通道数、图像尺寸、卷积核大小等。

//配置各网络层参数
network parse_network_cfg_custom(char *filename, int batch, int time_steps)
{
	// 从神经网络结构参数文件中读入所有神经网络层的结构参数,存储到sections中,
	// sections的每个node包含一层神经网络的所有结构参数
    list *sections = read_cfg(filename);
	// 获取sections的第一个节点,可以查看一下cfg/***.cfg文件,其实第一块参数(以[net]开头)不是某层神经网络的参数,
	// 而是关于整个网络的一些通用参数,比如学习率,衰减率,输入图像宽高,batch大小等,
	// 具体的关于某个网络层的参数是从第二块开始的,如[convolutional],[maxpool]...,
	// 这些层并没有编号,只说明了层的属性,但层的参数都是按顺序在文件中排好的,读入时,
	// sections链表上的顺序就是文件中的排列顺序。
    node *n = sections->front;
    if(!n) error("Config file has no sections");
	// 创建网络结构并动态分配内存:输入网络层数为sections->size - 1,sections的第一段不是网络层,而是通用网络参数
    network net = make_network(sections->size - 1);
	// 所用显卡的卡号(gpu_index在cuda.c中用extern关键字声明)
	// 在调用parse_network_cfg()之前,使用了cuda_set_device()设置了gpu_index的值号为当前活跃GPU卡号
    net.gpu_index = gpu_index;
	// size_params结构体元素不含指针变量
    size_params params;

    if (batch > 0) params.train = 0;    // allocates memory for Detection only
    else params.train = 1;              // allocates memory for Detection & Training

    section *s = (section *)n->val;
    list *options = s->options;
    if(!is_network(s)) error("First section must be [net] or [network]");
    parse_net_options(options, &net);

#ifdef GPU
    printf("net.optimized_memory = %d \n", net.optimized_memory);
    if (net.optimized_memory >= 2 && params.train) {
        pre_allocate_pinned_memory((size_t)1024 * 1024 * 1024 * 8);   // pre-allocate 8 GB CPU-RAM for pinned memory
    }
#endif  // GPU

    params.h = net.h;
    params.w = net.w;
    params.c = net.c;
    params.inputs = net.inputs;
    if (batch > 0) net.batch = batch;
    if (time_steps > 0) net.time_steps = time_steps;
    if (net.batch < 1) net.batch = 1;
    if (net.time_steps < 1) net.time_steps = 1;
    if (net.batch < net.time_steps) net.batch = net.time_steps;
    params.batch = net.batch;
    params.time_steps = net.time_steps;
    params.net = net;
    printf("mini_batch = %d, batch = %d, time_steps = %d, train = %d \n", net.batch, net.batch * net.subdivisions, net.time_steps, params.train);

    int avg_outputs = 0;
    float bflops = 0;
    size_t workspace_size = 0;
    size_t max_inputs = 0;
    size_t max_outputs = 0;
    n = n->next;
    int count = 0;
    free_section(s);

	// 此处stderr不是错误提示,而是输出结果提示,提示网络结构
    fprintf(stderr, "   layer   filters  size/strd(dil)      input                output\n");
    while(n){
        params.index = count;
        fprintf(stderr, "%4d ", count);
        s = (section *)n->val;
        options = s->options;
		// 定义网络层
        layer l = { (LAYER_TYPE)0 };
		// 获取网络层的类别

        LAYER_TYPE lt = string_to_layer_type(s->type);
		
		//通过读取网络类型,从而配置各网络层的参数
        if(lt == CONVOLUTIONAL){//yolov3-tiny  卷积层  13层
            l = parse_convolutional(options, params);
        }else if(lt == LOCAL){
            l = parse_local(options, params);
        }else if(lt == ACTIVE){
            l = parse_activation(options, params);
        }else if(lt == RNN){
            l = parse_rnn(options, params);
        }else if(lt == GRU){
            l = parse_gru(options, params);
        }else if(lt == LSTM){
            l = parse_lstm(options, params);
        }else if (lt == CONV_LSTM) {
            l = parse_conv_lstm(options, params);
        }else if(lt == CRNN){
            l = parse_crnn(options, params);
        }else if(lt == CONNECTED){
            l = parse_connected(options, params);
        }else if(lt == CROP){
            l = parse_crop(options, params);
        }else if(lt == COST){
            l = parse_cost(options, params);
            l.keep_delta_gpu = 1;
        }else if(lt == REGION){
            l = parse_region(options, params);
            l.keep_delta_gpu = 1;
        }else if (lt == YOLO) {//yolov3-tiny YOLO层  两层
            l = parse_yolo(options, params);
            l.keep_delta_gpu = 1;
        }else if (lt == GAUSSIAN_YOLO) {
            l = parse_gaussian_yolo(options, params);
            l.keep_delta_gpu = 1;
        }else if(lt == DETECTION){
            l = parse_detection(options, params);
        }else if(lt == SOFTMAX){
            l = parse_softmax(options, params);
            net.hierarchy = l.softmax_tree;
            l.keep_delta_gpu = 1;
        }else if(lt == NORMALIZATION){
            l = parse_normalization(options, params);
        }else if(lt == BATCHNORM){
            l = parse_batchnorm(options, params);
        }else if(lt == MAXPOOL){//yolov3-tiny 池化层 maxpool  6层
            l = parse_maxpool(options, params);
        }else if (lt == LOCAL_AVGPOOL) {
            l = parse_local_avgpool(options, params);
        }else if(lt == REORG){
            l = parse_reorg(options, params);        }
        else if (lt == REORG_OLD) {
            l = parse_reorg_old(options, params);
        }else if(lt == AVGPOOL){
            l = parse_avgpool(options, params);
        }else if(lt == ROUTE){//yolov3-tiny 路由层 2层
            l = parse_route(options, params);
            int k;
            for (k = 0; k < l.n; ++k) {
                net.layers[l.input_layers[k]].use_bin_output = 0;
                net.layers[l.input_layers[k]].keep_delta_gpu = 1;
            }
        }else if (lt == UPSAMPLE) {//yolov3-tiny 上采样层 1层
            l = parse_upsample(options, params, net);
        }else if(lt == SHORTCUT){
            l = parse_shortcut(options, params, net);
            net.layers[count - 1].use_bin_output = 0;
            net.layers[l.index].use_bin_output = 0;
            net.layers[l.index].keep_delta_gpu = 1;
        }else if (lt == SCALE_CHANNELS) {
            l = parse_scale_channels(options, params, net);
            net.layers[count - 1].use_bin_output = 0;
            net.layers[l.index].use_bin_output = 0;
            net.layers[l.index].keep_delta_gpu = 1;
        }
        else if (lt == SAM) {
            l = parse_sam(options, params, net);
            net.layers[count - 1].use_bin_output = 0;
            net.layers[l.index].use_bin_output = 0;
            net.layers[l.index].keep_delta_gpu = 1;
        }else if(lt == DROPOUT){
            l = parse_dropout(options, params);
            l.output = net.layers[count-1].output;
            l.delta = net.layers[count-1].delta;
            .........

下载权重文件

在parser.c的load_weights_upto中,根据卷积层的网络配置,开始下载读取各层的权重文件。

//读取权重文件函数
void load_weights_upto(network *net, char *filename, int cutoff)//cutoff = net->n
{
#ifdef GPU
    if(net->gpu_index >= 0){
        cuda_set_device(net->gpu_index);
    }
#endif
    fprintf(stderr, "Loading weights from %s...\n", filename);
    fflush(stdout);
    FILE *fp = fopen(filename, "rb");
    if(!fp) file_error(filename);

    int major;
    int minor;
    int revision;
    fread(&major, sizeof(int), 1, fp);//读取一个4字节的数据
    fread(&minor, sizeof(int), 1, fp);//读取一个4字节的数据
    fread(&revision, sizeof(int), 1, fp);//读取一个4字节的数据
	printf("the size of int in x64 is %d bytes,attention!!!\n", sizeof(int));//x86 x64: 4
	printf("major ,minor,revision of weight is %d, %d ,%d\n", major, minor, revision);//0.2.0
    if ((major * 10 + minor) >= 2) {//运行这一部分
        printf("\n seen 64");
        uint64_t iseen = 0;
        fread(&iseen, sizeof(uint64_t), 1, fp);//读取一个8字节的数据
		printf("the size of uint64_t is %d\n", sizeof(uint64_t));
        *net->seen = iseen;
    }
    else {
        printf("\n seen 32");
        uint32_t iseen = 0;
        fread(&iseen, sizeof(uint32_t), 1, fp);
        *net->seen = iseen;
    }
    *net->cur_iteration = get_current_batch(*net);
    printf(", trained: %.0f K-images (%.0f Kilo-batches_64) \n", (float)(*net->seen / 1000), (float)(*net->seen / 64000));
    int transpose = (major > 1000) || (minor > 1000);

    int i;
    for(i = 0; i < net->n && i < cutoff; ++i){//cutoff = net->n
        layer l = net->layers[i];
        if (l.dontload) continue;//always 0		跳过之后的循环体,直接运行++i
        if(l.type == CONVOLUTIONAL && l.share_layer == NULL){ //只运行这一个分支的代码
            load_convolutional_weights(l, fp);
			//printf("network layer [%d] is CONVOLUTIONAL \n",i);
        }
        .......

在读取yolov3-tiny各层权重文件前,先读取4个和训练有关的参数:major,minor, revision和iseen。在前向传播的工程当中,并没有实际的应用。

parser.c中的load_convolutional_weights函数,具体执行对yolov3-tiny权重文件的下载,包括节点参数weight,偏置参数bias和批量归一化参数BN。

void load_convolutional_weights(layer l, FILE *fp)
{
	static int flipped_num;
    if(l.binary){
        //load_convolutional_weights_binary(l, fp);
        //return;
    }
    int num = l.nweights;
	//int num = l.n*l.c*l.size*l.size;//l.n 输出的层数 l.c输入的层数 
    int read_bytes;
    read_bytes = fread(l.biases, sizeof(float), l.n, fp);//读取偏置参数 l.n个float数据
    if (read_bytes > 0 && read_bytes < l.n) printf("\n Warning: Unexpected end of wights-file! l.biases - l.index = %d \n", l.index);
    //fread(l.weights, sizeof(float), num, fp); // as in connected layer
    if (l.batch_normalize && (!l.dontloadscales)){
        read_bytes = fread(l.scales, sizeof(float), l.n, fp);//读取batch normalize 参数  l.n个float数据
        if (read_bytes > 0 && read_bytes < l.n) printf("\n Warning: Unexpected end of wights-file! l.scales - l.index = %d \n", l.index);
        read_bytes = fread(l.rolling_mean, sizeof(float), l.n, fp);//读取batch normalize 参数  l.n个float数据
        if (read_bytes > 0 && read_bytes < l.n) printf("\n Warning: Unexpected end of wights-file! l.rolling_mean - l.index = %d \n", l.index);
        read_bytes = fread(l.rolling_variance, sizeof(float), l.n, fp);//读取batch normalize 参数  l.n个float数据
        if (read_bytes > 0 && read_bytes < l.n) printf("\n Warning: Unexpected end of wights-file! l.rolling_variance - l.index = %d \n", l.index);

将权重参数批量归一化

yolov3-tiny每个卷积层之后,激活函数之前,都要对结果进行Batch Normalization:在这里插入图片描述
由于BN层和卷积操作都是线性的,将权重文件进行批量归一化,可以代替卷积层之后的BN层:
在这里插入图片描述
在network.c的fuse_conv_batchnorm函数中实现权重文件和BN层的合并。

void fuse_conv_batchnorm(network net)
{
    int j;
    for (j = 0; j < net.n; ++j) {
        layer *l = &net.layers[j];
		    // printf("the %d layer batch_normalize is %d,   groups is %d \n", j, l->batch_normalize, l->groups);
        if (l->type == CONVOLUTIONAL) { //只运行这一分支   合并卷积层和batch_normal
             //printf(" Merges Convolutional-%d and batch_norm \n", j);

            if (l->share_layer != NULL) {//l->share_layer always is 0,不运行这个分支
                l->batch_normalize = 0;
            }

            if (l->batch_normalize) {//#15,22层卷积,卷积之后没有batch normalize,其他都要运行这一分支
                int f;
                for (f = 0; f < l->n; ++f)//该层神经网络 1->n 个输出层权重
                {
                    l->biases[f] = l->biases[f] - (double)l->scales[f] * l->rolling_mean[f] / (sqrt((double)l->rolling_variance[f] + .00001));

                    const size_t filter_size = l->size*l->size*l->c / l->groups;//kernel_size * kernel_size * c/分组  l->groups存在于卷积层always is 1
                    int i;
                    for (i = 0; i < filter_size; ++i) {
                        int w_index = f*filter_size + i;

                        l->weights[w_index] = (double)l->weights[w_index] * l->scales[f] / (sqrt((double)l->rolling_variance[f] + .00001));
                    }
                }

                free_convolutional_batchnorm(l);//no use
                l->batch_normalize = 0;
                ......

输入图像

yolov3-tiny输入神经网络的图像尺寸为416×416,对不符合该尺寸的图像,要进行裁剪。在image.c的resize_image函数中完成。这个可以说是整个yolo算法对输入图像唯一进行预处理的地方了。这也是yolo算法在工程应用中极好的地方,没有那么多类似于降噪、滤波之类的预处理,直接送到网络里就完事了。

//im:输入图片  w:416 h:416
//函数作用:将输入图片热size到416x416的尺寸,基本按照缩放/扩大的策略
image resize_image(image im, int w, int h)
{
    if (im.w == w && im.h == h) return copy_image(im);

    image resized = make_image(w, h, im.c);//416 x 416 x 3空的地址空间
    image part = make_image(w, im.h, im.c);//416 x im.h x im.c空的地址空间
    int r, c, k;
    float w_scale = (float)(im.w - 1) / (w - 1);//宽度缩放因子
    float h_scale = (float)(im.h - 1) / (h - 1);//高度缩放因子
    for(k = 0; k < im.c; ++k){
        for(r = 0; r < im.h; ++r){
            for(c = 0; c < w; ++c){//416
                float val = 0;
                if(c == w-1 || im.w == 1){//c =415 最后一列
                    val = get_pixel(im, im.w-1, r, k);//取原图片最后一列的像素
                } else {
                    float sx = c*w_scale;
                    int ix = (int) sx;
                    float dx = sx - ix;
                    val = (1 - dx) * get_pixel(im, ix, r, k) + dx * get_pixel(im, ix+1, r, k);
                }
                set_pixel(part, c, r, k, val);
            }
        }
    }
    for(k = 0; k < im.c; ++k){
        for(r = 0; r < h; ++r){
            float sy = r*h_scale;
            int iy = (int) sy;
            float dy = sy - iy;
            for(c = 0; c < w; ++c){
                float val = (1-dy) * get_pixel(part, c, iy, k);
                set_pixel(resized, c, r, k, val);
            }
            if(r == h-1 || im.h == 1) continue;
            for(c = 0; c < w; ++c){
                float val = dy * get_pixel(part, c, iy+1, k);
                add_pixel(resized, c, r, k, val);
            }
        }
    }

    free_image(part);
    return resized;
}

前向传播网络

network.c中的forward_network函数是整个神经网络的核心部分,各层的网络都在函数指针l.forward(l, state)中完成。

void forward_network(network net, network_state state)
{
    state.workspace = net.workspace;
    int i;
	   /// 遍历所有层,从第一层到最后一层,逐层进行前向传播(网络总共有net.n层)
    for(i = 0; i < net.n; ++i){		  
        state.index = i;/// 置网络当前活跃层为当前层,即第i层		  
        layer l = net.layers[i];/// 获取当前层		  
        if(l.delta && state.train){//不执行此分支的代码
			/// 如果当前层的l.delta已经动态分配了内存,则调用fill_cpu()函数,将其所有元素的值初始化为0			   
            scal_cpu(l.outputs * l.batch, 0, l.delta, 1);/// 第一个参数为l.delta的元素个数,第二个参数为初始化值,为0
			printf("forward_network scal_cpu of %d layer done!\n ", i);
        }
           //double time = get_time_point();
		l.forward(l, state);//进行卷积运算,激活函数,池化运算/
		   //if layer_type = convolutional ;   l.forward = forward_convolutional_layer;
		   //if layer_type = maxpool           l.forward = forward_maxpool_layer;
		   //if layer_type = yolo              l.forward = forward_yolo_layer;
		   //if layer_type = ROUTE             l.forward = forward_route_layer;其实就是数据的复制和搬移
		   //if layer_type = upsample          l.forward = forward_upsample_layer;;		  
           //printf("%d - Predicted in %lf milli-seconds.\n", i, ((double)get_time_point() - time) / 1000);
		   /// 完成某一层的推理时,置网络的输入为当前层的输出(这将成为下一层网络的输入),要注意的是,此处是直接更改指针变量net.input本身的值,
		   /// 也就是此处是通过改变指针net.input所指的地址来改变其中所存内容的值,并不是直接改变其所指的内容而指针所指的地址没变,
		   /// 所以在退出forward_network()函数后,其对net.input的改变都将失效,net.input将回到进入forward_network()之前时的值。	
		   ......

卷积层[convolution]

卷积层在convolutional_layer.c中的forward_convolutional_layer函数实现。

void forward_convolutional_layer(convolutional_layer l, network_state state)
{
    
	int out_h = convolutional_out_height(l);//获得本层卷积层输出特征图的高、宽
    int out_w = convolutional_out_width(l);
    int i, j;
	
	// l.outputs = l.out_h * l.out_w * l.out_c在make各网络层函数中赋值(比如make_convolutional_layer()),
	// 对应每张输入图片的所有输出特征图的总元素个数(每张输入图片会得到n也即l.out_c张特征图)
	// 初始化输出l.output全为0.0;输入l.outputs*l.batch为输出的总元素个数,其中l.outputs为batch
	// 中一个输入对应的输出的所有元素的个数,l.batch为一个batch输入包含的图片张数;0表示初始化所有输出为0;
    fill_cpu(l.outputs*l.batch, 0, l.output, 1);//将地址l.output,l.outputs*l.batch个float地址空间的数据初始化0
    .......

作者在进行卷积运算前,将输入特征图进行重新排序:


```c
void im2col_cpu(float* data_im,
     int channels,  int height,  int width,
     int ksize,  int stride, int pad, float* data_col)
{
    int c,h,w;
	// 计算该层神经网络的输出图像尺寸(其实没有必要再次计算的,因为在构建卷积层时,make_convolutional_layer()函数
	// 已经调用convolutional_out_width(),convolutional_out_height()函数求取了这两个参数,
	// 此处直接使用l.out_h,l.out_w即可,函数参数只要传入该层网络指针就可了,没必要弄这么多参数)
    int height_col = (height + 2*pad - ksize) / stride + 1;
    int width_col = (width + 2*pad - ksize) / stride + 1;
	
	/// 卷积核大小:ksize*ksize是一个卷积核的大小,之所以乘以通道数channels,是因为输入图像有多通道,每个卷积核在做卷积时,
	/// 是同时对同一位置处多通道的图像进行卷积运算,这里为了实现这一目的,将三通道上的卷积核并在一起以便进行计算,因此卷积核
	/// 实际上并不是二维的,而是三维的,比如对于3通道图像,卷积核尺寸为3*3,该卷积核将同时作用于三通道图像上,这样并起来就得
	/// 到含有27个元素的卷积核,且这27个元素都是独立的需要训练的参数。所以在计算训练参数个数时,一定要注意每一个卷积核的实际
	/// 训练参数需要乘以输入通道数。
    int channels_col = channels * ksize * ksize;//输入通道
	// 外循环次数为一个卷积核的尺寸数,循环次数即为最终得到的data_col的总行数
    for (c = 0; c < channels_col; ++c) {

		//行,列偏置都是对应着本次循环要操作的输出位置的像素而言的,通道偏置,是该位置像素所在的输出通道的绝对位置(通道数)

		// 列偏移,卷积核是一个二维矩阵,并按行存储在一维数组中,利用求余运算获取对应在卷积核中的列数,比如对于
		// 3*3的卷积核(3通道),当c=0时,显然在第一列,当c=5时,显然在第2列,当c=9时,在第二通道上的卷积核的第一列,
		// 当c=26时,在第三列(第三输入通道上)
        int w_offset = c % ksize;//0,1,2
		// 行偏移,卷积核是一个二维的矩阵,且是按行(卷积核所有行并成一行)存储在一维数组中的,
		// 比如对于3*3的卷积核,处理3通道的图像,那么一个卷积核具有27个元素,每9个元素对应一个通道上的卷积核(互为一样),
		// 每当c为3的倍数,就意味着卷积核换了一行,h_offset取值为0,1,2,对应3*3卷积核中的第1, 2, 3行
        int h_offset = (c / ksize) % ksize;//0,1,2
		// 通道偏移,channels_col是多通道的卷积核并在一起的,比如对于3通道,3*3卷积核,每过9个元素就要换一通道数,
		// 当c=0~8时,c_im=0;c=9~17时,c_im=1;c=18~26时,c_im=2,操作对象是排序后的像素位置
        int c_im = c / ksize / ksize;
		// 中循环次数等于该层输出图像行数height_col,说明data_col中的每一行存储了一张特征图,这张特征图又是按行存储在data_col中的某行中
        for (h = 0; h < height_col; ++h) {
			// 内循环等于该层输出图像列数width_col,说明最终得到的data_col总有channels_col行,height_col*width_col列
            for (w = 0; w < width_col; ++w) {
				// 由上面可知,对于3*3的卷积核,行偏置h_offset取值为0,1,2,当h_offset=0时,会提取出所有与卷积核第一行元素进行运算的像素,
				// 依次类推;加上h*stride是对卷积核进行行移位操作,比如卷积核从图像(0,0)位置开始做卷积,那么最先开始涉及(0,0)~(3,3)
				// 之间的像素值,若stride=2,那么卷积核进行一次行移位时,下一行的卷积操作是从元素(2,0)(2为图像行号,0为列号)开始
                int im_row = h_offset + h * stride;//yolov3-tiny stride = 1
				// 对于3*3的卷积核,w_offset取值也为0,1,2,当w_offset取1时,会提取出所有与卷积核中第2列元素进行运算的像素,
				// 实际在做卷积操作时,卷积核对图像逐行扫描做卷积,加上w*stride就是为了做列移位,
				// 比如前一次卷积其实像素元素为(0,0),若stride=2,那么下次卷积元素起始像素位置为(0,2)(0为行号,2为列号)
                int im_col = w_offset + w * stride;
				// col_index为重排后图像中的像素索引,等于c * height_col * width_col + h * width_col +w(还是按行存储,所有通道再并成一行),
				// 对应第c通道,h行,w列的元素
                int col_index = (c * height_col + h) * width_col + w;//将重排后的图片像素,按照左上->右下的顺序,计算一维索引

				//im_col + width*im_row +  width*height*channel 重排前的特征图在内存中的位置索引
				// im2col_get_pixel函数获取输入图像data_im中第c_im通道,im_row,im_col的像素值并赋值给重排后的图像,
				// height和width为输入图像data_im的真实高、宽,pad为四周补0的长度(注意im_row,im_col是补0之后的行列号,
				// 不是真实输入图像中的行列号,因此需要减去pad获取真实的行列号)
                data_col[col_index] = im2col_get_pixel(data_im, height, width, channels,
                        im_row, im_col, c_im, pad);
				// return data_im[im_col + width*im_row +  width*height*channel)];
            }
        }
    }
}

通过gemm进行卷积乘加操作,通过add_bias添加偏置。

//进行卷积的乘加运算,没有bias偏置参数参与运算;
gemm(0, 0, m, n, k, 1, a, k, b, n, 1, c, n);

add_bias(l.output, l.biases, l.batch, l.n, out_h*out_w);//每个输出特征图的元素都加上对应通道的偏置参数

池化层[maxpool]

maxpool_layer.c中的forward_maxpool_layer函数完成池化操作。yolov3-tiny保留了池化层,并使用最大值池化,将尺寸为2×2的核中最大值保留下来。

void forward_maxpool_layer_avx(float *src, float *dst, int *indexes, int size, int w, int h, int out_w, int out_h, int c,
    int pad, int stride, int batch)
{

    const int w_offset = -pad / 2;
    const int h_offset = -pad / 2;
    int b, k;

    for (b = 0; b < batch; ++b) {
		// 对于每张输入图片,将得到通道数一样的输出图,以输出图为基准,按输出图通道,行,列依次遍历
		// (这对应图像在l.output的存储方式,每张图片按行铺排成一大行,然后图片与图片之间再并成一行)。
		// 以输出图为基准进行遍历,最终循环的总次数刚好覆盖池化核在输入图片不同位置进行池化操作。
        #pragma omp parallel for
        for (k = 0; k < c; ++k) {
            int i, j, m, n;
            for (i = 0; i < out_h; ++i) {
                //for (j = 0; j < out_w; ++j) {
                j = 0;
                for (; j < out_w; ++j) {
					// out_index为输出图中的索引
                    int out_index = j + out_w*(i + out_h*(k + c*b));//j + out_w * i + out_w * iout_h * k
                    float max = -FLT_MAX;// FLT_MAX为c语言中float.h定义的对大浮点数,此处初始化最大元素值为最小浮点数
                    int max_i = -1;// 最大元素值的索引初始化为-1
                    // 下面两个循环回到了输入图片,计算得到的cur_h以及cur_w都是在当前层所有输入元素的索引,内外循环的目的是找寻输入图像中,
                    // 以(h_offset + i*l.stride, w_offset + j*l.stride)为左上起点,尺寸为l.size池化区域中的最大元素值max及其在所有输入元素中的索引max_i
                    for (n = 0; n < size; ++n) {//2
                        for (m = 0; m < size; ++m) {//2
                            // cur_h,cur_w是在所有输入图像中第k通道中的cur_h行与cur_w列,index是在所有输入图像元素中的总索引。
                            // 为什么这里少一层对输入通道数的遍历循环呢?因为对于最大池化层来说输入与输出通道数是一样的,并在上面的通道数循环了!
                            int cur_h = h_offset + i*stride + n;
                            int cur_w = w_offset + j*stride + m;
                            int index = cur_w + w*(cur_h + h*(k + b*c));
							// 边界检查:正常情况下,是不会越界的,但是如果有补0操作,就会越界了,这里的处理方式是直接让这些元素值为-FLT_MAX
							// (注意虽然称之为补0操作,但实际不是补0),总之,这些补的元素永远不会充当最大元素值。
                            int valid = (cur_h >= 0 && cur_h < h &&
                                cur_w >= 0 && cur_w < w);
                            float val = (valid != 0) ? src[index] : -FLT_MAX;
							// 记录这个池化区域中的最大的元素值及其在所有输入元素中的总索引
                            max_i = (val > max) ? index : max_i;
                            max = (val > max) ? val : max;
                        }
                    }
					// 由此得到最大池化层每一个输出元素值及其在所有输入元素中的总索引。
					// 为什么需要记录每个输出元素值对应在输入元素中的总索引呢?因为在下面的反向过程中需要用到,在计算当前最大池化层上一层网络的敏感度时,
					// 需要该索引明确当前层的每个元素究竟是取上一层输出(也即上前层输入)的哪一个元素的值,具体见下面backward_maxpool_layer()函数的注释。
                    dst[out_index] = max;
                    if (indexes) indexes[out_index] = max_i;
                }
            }
        }
    }
}

路由层[route]

yolov3-tiny中共有两层路由层。第17层路由层(从0层开始),其实直接将第13层网络的输出结果输入。第20层路由层,将第19层和第8层网络结果合并在一起,19层在前,8层在后。在route_layer.c中的forward_route_layer函数中实现。

void forward_route_layer(const route_layer l, network_state state)
{
    int i, j;
    int offset = 0;
    for(i = 0; i < l.n; ++i){//l.n:  卷积层:输出特征图通道数 路由层:有几层网络层输入本层  17层:1(路由第13层)   20:2(路由第19、8层)
        int index = l.input_layers[i];//输入本网络层的网络层的索引:如13,19,8
        float *input = state.net.layers[index].output;//输入等于 之前网络层索引值得输出(.output)
        int input_size = l.input_sizes[i];//输入的网络层的数据量
        int part_input_size = input_size / l.groups;//未分组
        for(j = 0; j < l.batch; ++j){
            //copy_cpu(input_size, input + j*input_size, 1, l.output + offset + j*l.outputs, 1);
			//从首地址input处复制input_size 个数据到 l.output中
            copy_cpu(part_input_size, input + j*input_size + part_input_size*l.group_id, 1, l.output + offset + j*l.outputs, 1);//l.group_id = 0
			//其实就是copy_cpu(part_input_size, input, 1, l.output + offset, 1);
        }
        //offset += input_size;
        offset += part_input_size;
    }
}

上采样层[upsample]

yolov3-tiny中第19层是上采样层,将18层13x13x128的输入特征图转变为26x26x128的输出特征图。在upsample_layer.c中的forward_upsample_layer函数中完成。

void upsample_cpu(float *in, int w, int h, int c, int batch, int stride, int forward, float scale, float *out)
{
	
    int i, j, k, b;
    for (b = 0; b < batch; ++b) {
        for (k = 0; k < c; ++k) {
            for (j = 0; j < h*stride; ++j) {
                for (i = 0; i < w*stride; ++i) {
                    int in_index = b*w*h*c + k*w*h + (j / stride)*w + i / stride;
                    int out_index = b*w*h*c*stride*stride + k*w*h*stride*stride + j*w*stride + i;
                    if (forward) out[out_index] = scale*in[in_index];
                    else in[in_index] += scale*out[out_index];
                }
            }
        }
    }
}

上采样效果:
在这里插入图片描述

输出层[yolo]

yolo层完成了对13x13x255和26x26x255输入特诊图的logistic逻辑回归计算。每个box的预测宽度和高度不参与逻辑回归,在yolo_layer.c中的forward_yolo_layer函数中完成。

//两个yolo层 只对数据进行了logistic处理,并没有预测box的位置
//将0-1通道(x,y) 4-84(confidence+class)计算logistic,三个prior(预测框都是这样)
void forward_yolo_layer(const layer l, network_state state)
{
    int i, j, b, t, n;
	//从state.input复制数据到l.output
    memcpy(l.output, state.input, l.outputs*l.batch * sizeof(float));

#ifndef GPU
	printf("yolo v3 tiny l.n and l.batch of yolo layer is %d and %d  \n ",l.n,l.batch);
    for (b = 0; b < l.batch; ++b) {//l.batch = 1
        for (n = 0; n < l.n; ++n) {//l.n:3(yolo层)mask 0,1,2  表示每个网络单元预测三个box?
			
			//printf("l.coords is %d in yolov3 tiny yolo layer ,l.scale_x_y is %f \n", l.coords, l.scale_x_y);
            // l.coords 坐标:0  l.classes分类数量:80   l.scale_x_y:1
			//l.w:输入特征图宽度 l.h输出特征图高度  
			int index = entry_index(l, b, n*l.w*l.h, 0);//index = n*l.w*l.h*(4 + l.classes + 1)
			
		   //起始地址为:l.output + index 个数为:2 * l.w*l.h  计算逻辑回归值,并保存
            activate_array(l.output + index, 2 * l.w*l.h, LOGISTIC);  // x,y,

			//起始地址为:l.output + index 个数为:2 * l.w*l.h  计算方式为:x = x*l.scale_x_y + -0.5*(l.scale_x_y - 1) 简化后:x = x
			//yolov3-tiny l.scale_x_y = 1  实际上该函数没有参与任何的运算   scal_add_cpu
            scal_add_cpu(2 * l.w*l.h, l.scale_x_y, -0.5*(l.scale_x_y - 1), l.output + index, 1);    // scale x,y
            
			//
			index = entry_index(l, b, n*l.w*l.h, 4);//index = n*l.w*l.h*(4 + l.classes + 1)+ 4*l.w*l.h
            
			//起始地址为:l.output + index,个数为:(1+80)*l.w*l.h   计算器其逻辑回归值
			activate_array(l.output + index, (1 + l.classes)*l.w*l.h, LOGISTIC);
        }
    }

预测结果统计[detection ]


//w:输入图像宽度640,不一定是416 h:输入图像高度424,不一定是416  thresh:图像置信度阈值0.25   hier:0.5
//map:0   relative:1  num:0   letter:0
//函数作用:统计两个yolo层中 置信度大于阈值的box个数,并对这个box初始化一段地址空间 dets
//根据网络来填充该地址空间dets:
//根据yolo层 计算满足置信度阈值要求的box相对的预测坐标、宽度和高度,并将结果保存在dets[count].bbox结构体中
//每个box有80个类别,有一个置信度,该类别对应的可能性prob:class概率*置信度
///舍弃prob小于阈值0.25的box
//将满足阈值的box个数保存到num中
detection *get_network_boxes(network *net, int w, int h, float thresh, float hier, int *map, int relative, int *num, int letter)
{
	//printf("w、h、thresh、hier and letter is %d 、%d 、%f 、%f and %d\n", w, h, thresh, hier, letter);

	//函数作用:统计两个yolo层中 置信度大于阈值的box个数,并对这个box初始化一段地址空间 dets
	//将满足阈值的box个数保存到num中
    detection *dets = make_network_boxes(net, thresh, num);
	
	//根据网络来填充该地址空间dets:
	//根据yolo层 计算满足置信度阈值要求的box相对的预测坐标、宽度和高度,并将结果保存在dets[count].bbox结构体中
	//每个box有80个类别,有一个置信度,该类别对应的可能性prob:class概率*置信度
	///舍弃prob小于阈值0.25的box
    fill_network_boxes(net, w, h, thresh, hier, map, relative, dets, letter);
    return dets;
}

使用make_network_boxes来创建预测信息的指针变量:

// thresh:  置信度阈值
//num: 0
//函数作用:统计置信度大于阈值的box个数,并对这个box初始化一段地址空间
detection *make_network_boxes(network *net, float thresh, int *num)
{
    layer l = net->layers[net->n - 1];//应该是神经网络最后一层 net->n:24 最后一层yolo层
	//printf(" net->n  of network is %d\n " ,(net->n));
    int i;
	// -thresh 0.25
	//yolo层:yolov3-tiny中共有两层
	//三个prior预测框,对每个预测框中,置信度大于thresh 0.25,记为一次,将次数进行累加,并输出
	//nboxes:即为要保留的box的个数 两个yolo层中的置信度个数一起累加
    int nboxes = num_detections(net, thresh);//-thresh 0.25

	if (num) {
		printf("nbox = %d \n", num);
		*num = nboxes;//不执行该语句
	}
    //申请内存,个数为nboxes,每个内存大小为:sizeof(detection)
    detection* dets = (detection*)xcalloc(nboxes, sizeof(detection));

	//遍历每个box,每个dets.prob申请80个float类型的内存:
	//dets.uc,申请4个float类型的空间:位置信息
    for (i = 0; i < nboxes; ++i) {
        dets[i].prob = (float*)xcalloc(l.classes, sizeof(float));
        // tx,ty,tw,th uncertainty
        dets[i].uc = (float*)xcalloc(4, sizeof(float)); // Gaussian_YOLOv3
        
		if (l.coords > 4) {//不执行这个分支 l.coords:0
            dets[i].mask = (float*)xcalloc(l.coords - 4, sizeof(float));
        }
    }
    return dets;
}

使用get_yolo_detections来统计两层yolo层的预测信息:

//w,h:640,424    netw, neth:416,416 thresh:图像置信度阈值0.25   hier:0.5
//map:0   relative:1    letter:0
//根据yolo层 计算满足置信度阈值要求的box相对的预测坐标、宽度和高度,并将结果保存在dets[count].bbox结构体中
//每个box有80个类别,有一个置信度,该类别对应的可能性prob:class概率*置信度
///舍弃prob小于阈值0.25的box
int get_yolo_detections(layer l, int w, int h, int netw, int neth, float thresh, int *map, int relative, detection *dets, int letter)
{
    printf("\n l.batch = %d, l.w = %d, l.h = %d, l.n = %d ,netw = %d, neth = %d \n", l.batch, l.w, l.h, l.n, netw, neth);
    int i,j,n;
    float *predictions = l.output;//yolo层的输出
    // This snippet below is not necessary
    // Need to comment it in order to batch processing >= 2 images
    //if (l.batch == 2) avg_flipped_yolo(l);
    int count = 0;

	//printf("yolo layer l.mask[0] is %d, l.mask[1] is %d, l.mask[2] is %d\n", l.mask[0], l.mask[1], l.mask[2]);
	//printf("yolo layer l.biases[l.mask[0]*2] is %f, l.biases[l.mask[1]*2] is %f, l.biases[l.mask[2]*2] is %f\n", l.biases[l.mask[0] * 2], l.biases[l.mask[1] * 2], l.biases[l.mask[2] * 2]);
	//遍历yolo层
    for (i = 0; i < l.w*l.h; ++i){//该yolo层输出特征图的宽度、高度:13x13 26x26
        int row = i / l.w;
        int col = i % l.w;
        for(n = 0; n < l.n; ++n){//yolo层,l.n = 3
			
            //obj_index:置信度层索引
            int obj_index  = entry_index(l, 0, n*l.w*l.h + i, 4);//obj_index  = n*l.w*l.h*(4+l.classes+1) + 4*l.w*l.h + i;
            float objectness = predictions[obj_index];//获得对应的置信度
            //if(objectness <= thresh) continue;    // incorrect behavior for Nan values
            
			if (objectness > thresh) {//只有置信度大于阈值才开始执行该分支
                //printf("\n objectness = %f, thresh = %f, i = %d, n = %d \n", objectness, thresh, i, n);
                
				//box_index:yolo层每个像素点有三个box,表示每个box的索引值
				int box_index = entry_index(l, 0, n*l.w*l.h + i, 0);//box_index = n*l.w*l.h*(4+l.classes+1)+ i;

				//l.biases->偏置参数起始地址    l.mask[n]:分别为3,4,5,0,1,2,biases偏置参数偏移量
				//根据yolo层 计算满足置信度阈值要求的box相对的预测坐标、宽度和高度,并将结果保存在dets[count].bbox结构体中
                dets[count].bbox = get_yolo_box(predictions, l.biases, l.mask[n], box_index, col, row, l.w, l.h, netw, neth, l.w*l.h);

				//获取对应的置信度,该置信度经过了logistic
                dets[count].objectness = objectness;

				//获得分类数:80(int类型)
                dets[count].classes = l.classes;
                for (j = 0; j < l.classes; ++j) {
					//80个类别,每个类别对应的概率,class_index为其所在层的索引
                    int class_index = entry_index(l, 0, n*l.w*l.h + i, 4 + 1 + j);//class_index  = n*l.w*l.h*(4+l.classes+1) + (4+1+j)*l.w*l.h + i;
                    //每个box有80个类别,有一个置信度,该类别对应的可能性prob:class概率*置信度
					float prob = objectness*predictions[class_index];
					
					//舍弃prob小于阈值0.25的box
                    dets[count].prob[j] = (prob > thresh) ? prob : 0;
                }
                ++count;
            }
        }
    }
    correct_yolo_boxes(dets, count, w, h, netw, neth, relative, letter);
    return count;
}

非极大值抑制[NMS]

//dets:box结构体 nboxes:满足阈值的box个数   l.classe:80    thresh=0.45f
//两个box,同一类别进行非极大值抑制,遍历
void do_nms_sort(detection *dets, int total, int classes, float thresh)
{
    int i, j, k;
    k = total - 1;
    for (i = 0; i <= k; ++i) {//box个数
        if (dets[i].objectness == 0) {//置信度==0  不执行该分支,理论上没有objectness = 0
			printf("there is no objectness == 0 !!! \n");
            detection swap = dets[i];
            dets[i] = dets[k];
            dets[k] = swap;
            --k;
            --i;
        }
    }
    total = k + 1;
	//同一类别进行比较
    for (k = 0; k < classes; ++k) {//80个        
        //box预测的类别
		for (i = 0; i < total; ++i) {//box个数
            dets[i].sort_class = k;
        }
		//函数作用:将prob较大的box排列到前面
        qsort(dets, total, sizeof(detection), nms_comparator_v3);
        for (i = 0; i < total; ++i) {//两个box,同一类别进行非极大值抑制
            //printf("  k = %d, \t i = %d \n", k, i);
            if (dets[i].prob[k] == 0) continue;
            box a = dets[i].bbox;
            for (j = i + 1; j < total;++j){
				box b = dets[j].bbox;
				if( box_iou(a, b) > thresh) dets[j].prob[k] = 0;
            }
        }
    }
}

FPGA—YOLO

device = “xcvu3p-ffvc1517-2-e”

参考:

https://thedatabus.io/convolver

https://github.com/sumanth-kalluri/cnn_hardware_acclerator_for_fpga

https://thedatabus.io/archive

1、网络结构分析:

YOLOv3 的剪枝量化:

https://github.com/coldlarry/YOLOv3-complete-pruning

说明:最后一层75输出通道

输出为单个通道:上述过程中,每一个卷积核的通道数量,必须要求与输入通道数量一致,因为要对每一个通道的像素值要进行卷积运算,所以每一个卷积核的通道数量必须要与输入通道数量保持一致

preview

多通道

输入 416*416*3 = 519,168 个值

conv0 : input_height=418 原因:这里是做1的填充,所以 最终输入是 416+2 =418

 conv_0_param={.kernel_dim=3,
                          .pad=1,
                          .input_channel=3,
                          .input_width=418,
                          .input_height=418,
                          .output_channel=16,
                          .output_width=416,
                          .output_height=416};

16*3个卷积核:一共16*3*9个卷积核参数 +16个bias == 》 432+16 个参数

输出: 416 *416* 16 个值,即2768896个输出

conv1 :

conv_1_param={.kernel_dim=3,
                          .pad=1,
                          .input_channel=16,
                          .input_width=210,
                          .input_height=210,
                          .output_channel=32,
                          .output_width=208,
                          .output_height=208};

一共 32*16个卷积核 4608个卷积核权重参数,36个bias参数

该层输出: 208 *208* 32 = 1384448

conv2 :

conv_2_param={.kernel_dim=3,
                          .pad=1,
                          .input_channel=32,
                          .input_width=106,
                          .input_height=106,
                          .output_channel=64,
                          .output_width=104,
                          .output_height=104};

一共 64 *32个卷积核 18432个卷积核权重参数,64个bias参数

该层输出: 104 *104*64 = 692224

conv3:

conv_3_param={.kernel_dim=3,
                          .pad=1,
                          .input_channel=64,
                          .input_width=54,
                          .input_height=54,
                          .output_channel=128,
                          .output_width=52,
                          .output_height=52};

一共 128*64个卷积核 73728个卷积核权重参数,128个bias参数

该层输出: 52 *52*128 = 346112

conv4: 256*128*9= 294912 个卷积核权重参数 ,256个偏置, 输出:26*26*256=173056

conv_4_param={.kernel_dim=3,
                          .pad=1,
                          .input_channel=128,
                          .input_width=28,
                          .input_height=28,
                          .output_channel=256,
                          .output_width=26,
                          .output_height=26};

conv5:256*512*9 = 1179648个卷积核权重 512 个bias, 输出:13*13*512=86528

conv6:512*1024*9=4718592个卷积核权重,1024个bias,输出:13*13*1024=173056

conv7:1024*256*1 =262144个卷积层权重,256个bias,输出:13*13*256=43264

conv8:256*512*9 =1179648 个卷积层权重 ,512个bias 输出:13*13*512=86528

上面的卷积层:后面紧跟一个batchnormal层

CONV 9 and 12 have no batch norm layer

conv9 : 最后一层:512*255*1=130560个卷积权重,255个bias, 13*13*255= 43095

conv10: 256*128*1=32768个权重 128个bias 输出13*13*128=21632

conv11: 384*256*3 =305280 个参数 256个bias 输出26*26*256=173056

conv12: 256*255*1=65280个权重 255个bias 输出26*26*255 = 172380

wight= [448,4608,36,18432,64,73728,128,294912,256,1179648,512,4718592,1024,262144,256,1179648 ,512,130560,255,32768,128,305280,256,65280,255]
out: [ 519168,2768896,1384448, 692224,346112,173056,86528,173056,43264,86528,43095,21632,173056,172380]

汇总:8269730个wight((八百多万) 输出结果中最大 2768896,最小21632

数据位宽:

在定点实现中,首先确定整数和小数部分使用了多少位是很重要的。对于线性量化,整数的位宽度与极值以及是否会发生溢出有关。另一方面,分数部分的长度影响量化误差。此外,量化的步长将影响数据的分布。对于不同的网络类型,应在比特宽度和网络精度之间进行彻底的权衡。

在定点实现中,首先确定整数和小数部分使用了多少位是很重要的。对于线性量化,整数的位宽度与极值以及是否会发生溢出有关。另一方面,分数部分的长度影响量化误差。此外,量化的步长将影响数据的分布。对于不同的网络类型,应在比特宽度和网络精度之间进行彻底的权衡。

就总比特而言,2的幂次更可取。为了确定合适的比特宽度,在包含5000幅图像的COCO2014-val5k数据集上进行了实验。下表总结了所有卷积层的数据分布。很明显,所有输入和输出的绝对值都小于128。这意味着在不造成任何溢出的情况下,为有符号整数部分分配8位就足够了。

尝试使用8位表示整数部分,另外8位表示小数。建立了软件定点仿真系统:

因此选择16bit存储权重以及16bit的数据权重,两者都由16位定点数字表示。对于卷积窗口的结果,采用32位以减少可能的溢出,并提供更高的累积精度。通过实施定点优化,数据宽度从32位压缩到16位,从而实现更高效的DMA传输。此外,使用定点表示为延迟和资源带来了实实在在的好处。

激活函数实现:

官方 float leaky_activate(float x){return (x>0) ? x : .1*x;}

使用动态定点量化时,激活函数 Leaky ReLU 也要进行相应的量化操作。当位宽 bw 为 16 时,使用与 0xccc 相乘和右移 15 位的定点运算来拟合与 0.1 相乘的操作。量化后的 Leaky ReLU 如下所示:
𝑓′(𝑥) = (𝑥 < 0)? (𝑥 ∗ 0xccc) ≫ 15: 𝑥

合并 Batch Normalisation

OLOv3 tiny中的大多数卷积层在激活前都会进行批量归一化。使用一些数学技术消除批次标准化是安全的。可以很容易地得到以下方程式。

w’0ij和b’j是新的权重和偏差。因为这种转换可以在运行时之前完成,浮点计算过程中的一些可忽略的精度损失外。合并批次归一化也会影响权重的分布。

二维卷积:

2D 卷积是一种采用权重内核(或窗口或矩阵)并使用这些权重(此矩阵中的值)以某种方式修改输入图像(或特征图或激活图)的操作,pad 默认补0操作

激活函数

在神经网络中的各种数学线性运算之间引入了激活函数,目的是为整个网络引入非线性。没有它,整个神经网络将简化为单个线性函数,我们都知道这样的函数不足以模拟任何复杂的东西,忘记像图像识别这样的东西。

文献中有多种用于此目的的函数,但最常用(也是第一个)使用的函数是 ReLu(整流线性单元)和 Tanh(双曲正切)函数。

池化函数

池化本质上是一种“下采样”操作,旨在减少数据在网络中传播时的参数数量和复杂性。此过程涉及在输入上运行“池化窗口”并使用某种算法减小输入的大小。到目前为止,最常见的算法是 Max – Pooling 和 Average – Pooling。

最大池化——在这种方法中,我们只保留落入池化窗口的所有值中的最大值,并丢弃其他值。

平均池化 – 在此方法中,计算池化窗口内所有元素的平均值,并保留该值而不是所有值。

upsaming 上采样:

torch.nn.functional.upsample

nearest 采样

yolov3 中的上采样层使用的是2*2上采样:

import torch
from torch import nn
input = torch.arange(1, 5, dtype=torch.float32).view(1, 1, 2, 2)
input


tensor([[[[1., 2.],
          [3., 4.]]]])
 

m = nn.Upsample(scale_factor=2, mode='nearest')
m(input)
返回:

tensor([[[[1., 1., 2., 2.],
          [1., 1., 2., 2.],
          [3., 3., 4., 4.],
          [3., 3., 4., 4.]]]])

最邻近插值算法

首先假设原图是一个像素大小为W*H的图片,缩放后的图片是一个像素大小为w*h的图片,这时候我们是已知原图中每个像素点上的像素值(即灰度值等)的(⚠️像素点对应像素值的坐标都是整数)。这个时候已知缩放后有一个像素点为(x,y),想要得到该像素点的像素值,那么就要根据缩放比例去查看其对应的原图的像素点的像素值,然后将该像素值赋值给该缩放后图片的像素点(x,y)

缩放公式为:

  • 根据横轴,即宽可得:X/x = W/w
  • 根据纵轴,即高可得:Y/y = H/h
  • 那么能够得到 f(X,Y)= f( W/w * x, H/h *y)

因此这个时候缩放后的图片像素点(x,y)的像素值就对应着原图像素点( W/w * x, H/h *y)的像素值

但是这个时候会出现一个问题就是因为缩放比例的原因,会导致像素点( W/w * x, H/h *y)中的值不是整数,那么就不知道应该对应的是哪个像素点的像素值

这个时候最邻近插值算法使用的方法就是四舍五入法,表示为[.],所以像素值f(x,y) = f( [W/w * x],  [H/h *y])

举个例子,如果原图为5*5,缩放后的图为3*3,那么缩放后的图的像素点(1,1)对应的就是原图中([5/3 * 1], [5/3 * 1]) = ([0.6], [0.6]) = (1,1) 像素点对应的像素值

这种方法的好处就是简单,但是坏处就是太过粗暴,会缺失精度,造成缩放后的图像灰度上的不连续,在变化地方可能出现明显锯齿状,如下图所示:(左原图,右缩放后)

累加层accumulate:

在yolo中存在累加操作:conv层 的输出与另外一个conv层输出加和:

yolo输出层

yolo在经过多个卷积和上采样之后最终得到的是2个个卷积结果(13*13*255,26*26*255),每一个卷积结果的长和宽分别是(13×13,26×26),深度信息是 [4(box信息)+1(物体判别信息)+80(classNum置信度)] *3(每个点上面计算几个anchor)

对于具有Gh×Gw×Nin输入的YoLO层,它将原始图像划分为 Gh×Gw 网格。通道数n等于(4 +1+C)×B,其中B表示在一个网格中可以检测到多少个对象,C表示对象类别的数量。在YOLOv3 tiny中,B设置为3,COCO数据集中有80个类(c=80).

因此,YLO层的输入具有255的恒定值,这进一步分为3组( 表示在一个网格中可以检测 3种对象)。在每组中,4个通道提供关于边界框的信息,1个通道表示对象分数,其余80个通道表示单个类分数。Yolo层在所有通道上使用Sigmod激活函数,除了表示边界框宽度和高度的通道以外。关键在于指定哪些通道应该通过sigmoid函数,哪些通道应该不被触及。解决方案是提供一个表,其中每个位都表示是否应转换通道。

最终得到的边框坐标值是bx,by,bw,bh即边界框bbox相对于feature map的位置和大小,是我们需要的预测输出坐标。但网络实际上的学习目标是tx,ty,tw,th这4个偏移量(offsets,其中tx,ty是预测的坐标偏移值,tw,th是尺度缩放,有了这4个offsets,自然可以根据图2的公式去求得真正需要的bx,by,bw,bh4个坐标。bx,by是预测框中心坐标,bw,bh是预测框的宽高

其中 \(C_{x} C_{y}\)是 feature map上anchor box的宽和高, \(t_{y}, t_{w} t_{z} t_{x}\)是 4个通道提供关于边界框的信息

Cx,Cy是feature map中grid cell的左上角坐标,在yolov3中每个grid cell在feature map中的宽和高均为1。如图3的情形时,这个bbox边界框的中心属于第二行第二列的grid cell,它的左上角坐标为(1,1),故Cx=1,Cy=1。Pw、Ph是预设的anchor box映射到feature map中的宽和高,在yolov3.cfg文件中的anchor box原本设定是相对于416*416坐标系下的坐标,代码中是把cfg中读取的坐标除以stride如32映射到feature map坐标系中。

yolov3.cfg文件中的anchor box :

anchors = 10,14,  23,27,  37,58,  81,82,  135,169,  344,319

问题:anchor box作用的详细描述。
解答:YOLO3为每种FPN预测特征图(13*13,26*26,52*52)设定3种anchor box,总共聚类出9种尺寸的anchor box。在COCO数据集这9个anchor box是:(10×13),(16×30),(33×23),(30×61),(62×45),(59×119),(116×90),(156×198),(373×326)。分配上,在最小的13*13特征图上由于其感受野最大故应用最大的anchor box (116×90),(156×198),(373×326),(这几个坐标是针对416*416下的,当然要除以32把尺度缩放到13*13下),适合检测较大的目标。中等的26*26特征图上由于其具有中等感受野故应用中等的anchor box (30×61),(62×45),(59×119),适合检测中等大小的目标。较大的52*52特征图上由于其具有较小的感受野故应用最小的anchor box(10×13),(16×30),(33×23),适合检测较小的目标。同Faster-Rcnn一样,特征图的每个像素(即每个grid)都会有对应的三个anchor box,如13*13特征图的每个grid都有三个anchor box (116×90),(156×198),(373×326)(这几个坐标需除以32缩放尺寸)。

卷积:

yolo卷积默认补0

主要思想是构建一个高度流水线的流式架构,其中处理模块不必在任何时间点停止。即卷积器的任何部分都不会等待任何其他部分完成其工作并提供结果。每个阶段在每个时钟周期的输入的不同部分上连续执行整个工作的一小部分。这不仅仅是该设计的一个特点,它是一种称为“流水线”的一般原则,广泛用于将大型计算过程分解为更小的步骤,并提高整个电路可以运行的最高频率。

  • 设计使用 MAC(乘法和累加)单元,旨在将这些操作映射到 FPGA 的 DSP 模块。实现这一点将使乘法和加法运算更快,并且消耗更少的功率,因为​​ DSP 模块是在硬宏中实现的。也就是说,DSP 模块已经以最有效的方式合成、放置和路由到 FPGA 设备上的硅片中,这与一般 IP 模块不同,后者只向您提供经过试验和测试的 RTL 代码并且可以合成随心所欲。
  • 卷积器只不过是一组 MAC 单元和一些移位寄存器,当它们提供正确的输入时,在固定数量的时钟周期后输出卷积的结果
  • 输入特征图(图像)的大小是一个维度NxN,我们的内核(过滤器/窗口)是维度KxK。显然可以理解,即K < N.
  • s表示窗口在特征图上移动的步幅值。
  • 我们还有一些信号,例如valid_convend_conv信号,它们告诉外界这个卷积器模块的输出是否有效。但是为什么卷积器首先会产生无效的结果呢?窗口在输入的特定行上完成移动后,它会继续环绕到下一行,从而创建无效输出。可以避免在环绕期间进行计算,但这需要我们停止流水线,这违反了我们的流式设计原则。因此,我们只是在信号的帮助下丢弃我们认为无效的输出valid conv

具体流程:

下一次从a1开始输入,输出 输出 = w0*a01+ w1*a2 + w2*a3 + w3*a5 + w4*a6 + w5*a7 + w6*a9 + w7*a10 + w8*a11

卷积动画

输出 = w0*a0 + w1*a1 + w2*a2 + w3*a4 + w4*a5 + w5*a6 + w6*a8 + w7*a9 + w8*a10

这种架构的优点:

  • 输入特征图只需发送一次(即存储的激活必须从内存中访问一次),从而大大减少了内存访问时间和存储需求。移位寄存器为先前访问的值创建一种本地缓存。
  • 可以使用此架构计算任何大小输入的卷积,而无需由于计算能力低而中断输入或将其临时存储在其他地方
  • 对于一些需要更大步幅的奇异架构,步幅值也可以更改。

代码:

import numpy as np 
from scipy import signal

ksize = 3      															#ksize x ksize convolution kernel
im_size = 4   															#im_size x im_size input activation map

def strideConv(arr, arr2, s): 											  #the function that performs the 2D convolution
    return signal.convolve2d(arr, arr2[::-1, ::-1], mode='valid')[::s, ::s]

kernel =  np.arange(0,ksize*ksize,1).reshape((ksize,ksize))                   #the kernel is a matrix of increasing numbers
act_map = np.arange(0,im_size*im_size,1).reshape((im_size,im_size))           #the activation map is a matrix of increasin numbers

conv = strideConv(act_map,kernel,1)
print(kernel)
print(act_map)
print(conv) 

单个加法:

module mac_manual #(
	parameter N = 16,
    parameter Q = 12
    )(
    input clk,sclr,ce,
    input [N-1:0] a,
    input [N-1:0] b,
    input [N-1:0] c,
    output reg [N-1:0] p
    );

always@(posedge clk,posedge sclr)
 begin
    if(sclr)
    begin
        p<=0;
    end
    else if(ce)
    begin
        p <= (a*b+c);                   //performs the multiply accumulate operation
    end
 end
endmodule


卷积中的偏置:
preview

多卷积核的偏置:同一个通道的卷积核共用一个偏置。每个卷积核卷积后的结果+bias

https://cs231n.github.io/assets/conv-demo/index.html

定点表示:

定点数字是小数点位置保持固定的数字,与数字所代表的值无关。与浮点数相比,这使得定点数更易于理解以及在硬件中实现。与浮点运算相比,定点运算使用的资源也少得多。当然,所有这些都需要权衡。浮点运算可以为特定位宽提供比定点运算更高的精度。当我们使用二进制点位于固定位置的数字时,我们会在量化噪声方面受到打击,尽管该数字表示幅度。

Verilog HDL语法

  1. 算数运算符:+ – * / %(取模)
  2. 赋值运算符 =(堵塞) <=(非堵塞)
  3. 关系运算符 > ,< , >= , <= , ==:逻辑相等, !=:逻辑不等,===:全等;!==:不全等 “===”和”!==”可以比较含有x和z的操作数,在模块的功能仿真中有着广泛的应用。
  4. 逻辑运算符: && || !
  5. 位运算符: 按位 & | !~ ,^按位异或 ,^~ 按位同或
  6. 条件运算符: ?: r=s?t:u
  7. 移位运算符: << >>
  8. 拼接运算符: {信号1的某几位,信号2的某几位,信号3的某几位….} 2{信号的某几位} 表示{}中的信号重复2次
  9. 缩减运算符: 单目运算符 &,|,~ : c= &b 指的是b的前一位&后一位,知道最终只剩下一个bit

Vector:

type [upper:lower] vector_name;

如果声明vector时候指定了vector方向 [8:0] 高到低,那么在使用该vector时,也要从高到低指定:[3:0] [4:1]

向量的字节顺序(或非正式地称为“方向”)是指最低有效位是具有较低的索引(较小的字节序,例如[3:0])还是具有较高的索引(较大的字节序,例如[[ 0:3])。在Verilog中,一旦以特定的字节序声明了向量,就必须始终以相同的方式使用它。例如,声明vec[0:3]时写是非法的。与字节序一致是一种好习惯,因为如果将不同字节序的向量一起分配或使用,则会发生奇怪的错误。

wire [7:0] w;         // 8-bit wire
reg  [4:1] x;         // 4-bit reg
output reg [0:0] y;   // 1-bit reg that is also an output port (this is still a vector)
input wire [3:-2] z;  // 6-bit wire input (negative ranges are allowed)
output [3:0] a;       // 4-bit output wire. Type is 'wire' unless specified otherwise.
wire [0:7] b;         // 8-bit wire where b[0] is the most-significant bit.
wire [2:0] a, c;   // Two vectors
assign a = 3'b101;  // a = 101
assign b = a;       // b =   1  implicitly-created wire
assign c = b;       // c = 001  <-- bug
my_module i1 (d,e); // d and e are implicitly one-bit wide if not declared.
                    // This could be a bug if the port was intended to be a vector.

Adding `default_nettype none would make the second line of code an error, which makes the bug more visible.
assign w = a;
takes the entire 4-bit vector a and assigns it to the entire 8-bit vector w (declarations are taken from above). If the lengths of the right and left sides don't match, it is zero-extended or truncated as appropriate.

wire 和 reg 的区别和使用:

wire表示直通,即输入有变化,输出马上无条件地反映(如与、非门的简单连接)。

reg表示一定要有触发,输出才会反映输入的状态。

reg相当于存储单元,wire相当于物理连线。reg表示一定要有触发,没有输入的时候可以保持原来的值,但不直接实际的硬件电路对应。

      两者的区别是:寄存器型数据保持最后一次的赋值,而线型数据需要持续的驱动。wire使用在连续赋值语句中,而reg使用在过程赋值语句(initial ,always)中。wire若无驱动器连接,其值为z,reg默认初始值为不定值 x 。

     在连续赋值语句中,表达式右侧的计算结果可以立即更新表达式的左侧。在理解上,相当于一个逻辑之后直接连了一条线,这个逻辑对应于表达式的右侧,而这条线就对应于wire。在过程赋值语句中,表达式右侧的计算结果在某种条件的触发下放到一个变量当中,而这个变量可以声明成reg类型的。根据触发条件的不同,过程赋值语句可以建模不同的硬件结构:如果这个条件是时钟的上升沿或下降沿,那么这个硬件模型就是一个触发器;如果这个条件是某一信号的高电平或低电平,那么这个硬件模型就是一个锁存器;如果这个条件是赋值语句右侧任意操作数的变化,那么这个硬件模型就是一个组合逻辑。

      对组合逻辑输出变量,可以直接用assign。即如果不指定为reg类型,那么就默认为1位wire类型,故无需指定1位wire类型的变量。当然专门指定出wire类型,可能是多位或为使程序易读。wire只能被assign连续赋值,reg只能在initial和always中赋值。

      输入端口可以由wire/reg驱动,但输入端口只能是wire;输出端口可以是wire/reg类型,输出端口只能驱动wire;若输出端口在过程块中赋值则为reg型,若在过程块外赋值则为net型(wire/tri)。用关键词inout声明一个双向端口, inout端口不能声明为reg类型,只能是wire类型。

      默认信号是wire类型,reg类型要申明。这里所说的默认是指输出信号申明成output时为wire。如果是模块内部信号,必须申明成wire或者reg.

      对于always语句而言,赋值要申明成reg,连续赋值assign的时候要用wire。

模块调用时 信号类型确定方法总结如下:

•信号可以分为端口信号和内部信号。出现在端口列表中的信号是端口信号,其它的信号为内部信号。

•对于端口信号,输入端口只能是net类型。输出端口可以是net类型,也可以是register类型。若输出端口在过程块中赋值则为register类型;若在过程块外赋值(包括实例化语句),则为net类型。

•内部信号类型与输出端口相同,可以是net或register类型。判断方法也与输出端口相同。若在过程块中赋值,则为register类型;若在过程块外赋值,则为net类型。

•若信号既需要在过程块中赋值,又需要在过程块外赋值。这种情况是有可能出现的,如决断信号。这时需要一个中间信号转换。

下面所列是常出的错误及相应的错误信息(error message)

•用过程语句给一个net类型的或忘记声明类型的信号赋值。

           信息:illegal …… assignment.

•将实例的输出连接到声明为register类型的信号上。

           信息:<name> has illegal output port specification.

•将模块的输入信号声明为register类型。

case 、casex和casez的区别

用法:

case(in)
 2'b00:out = 0;
 2'b01:out = 1;
 2'b10:out = 2;
 2'b11:out = 3;
endcase

casex: 该语句不考虑高阻值z以及不定值x,即在表达式进行 比较时,不会比较x、z所在位的状态:in = 001 c = xx1 那么 in在case语句中等于c,因为不考虑x位,只需要比较最后一位 1=1


//要实现 如果in的第一个1出现的位置 in == 000 则out = 0  in = 101则out =1  in =111  则out仍是1
case(in)
 2'bxx1:out = 3;
 2'bx1x:out = 2;
 2'b1xx:out = 1;
 2'b000:out = 0
endcase

同理 casez 该语句不考虑高阻值z , 即在表达式进行 比较时,不会比较z所在位的状态

AXI4协议

ZYNQ将高性能ARM Cotex-A系列处理器与高性能FPGA在单芯片内紧密结合,为设计带来了如减小体积和功耗、降低设计风险,增加设计灵活性等诸多优点。在将不同工艺特征的处理器与FPGA融合在一个芯片上之后,片内处理器与FPGA之间的互联通路就成了ZYNQ芯片设计的重中之重。如果Cotex-A9与FPGA之间的数据交互成为瓶颈,那么处理器与FPGA结合的性能优势就不能发挥出来。

AXI的英文全称是Advanced eXtensible Interface,即高级可扩展接口,它是ARM公司所提出的AMBA(Advanced Microcontroller Bus Architecture)协议的一部分

AXI协议就是描述主设备和从设备之间的数据传输方式,在该协议中,主设备和从设备之间通过握手信号建立连接。

AXI协议是一种高性能、高带宽、低延迟的片内总线,具有如下特点:
1、总线的地址/控制和数据通道是分离的;
2、支持不对齐的数据传输;
3、支持突发传输,突发传输过程中只需要首地址;
4、具有分离的读/写数据通道;
5、支持显着传输访问和乱序访问;
6、更加容易进行时序收敛。

在数字电路中只能传输二进制数0和1,因此可能需要一组信号才能高效地传输信息,这一组信号就组成了接口。AXI4协议支持以下三种类型的接口:
1、 AXI4:高性能存储映射接口。
2、 AXI4-Lite:简化版的AXI4接口,用于较少数据量的存储映射通信。
3、 AXI4-Stream:用于高速数据流传输,非存储映射接口。

在这里我们首先解释一下存储映射(Meamory Map)这一概念。如果一个协议是存储映射的,那么主机所发出的会话(无论读或写)就会标明一个地址。这个地址对应于系统存储空间中的一个地址,表明是针对该存储空间的读写操作。

AXI4协议支持突发传输,主要用于处理器访问存储器等需要指定地址的高速数据传输场景。AXI-Lite为外设提供单个数据传输,主要用于访问一些低速外设中的寄存器。而AXI-Stream接口则像FIFO一样,数据传输时不需要地址,在主从设备之间直接连续读写数据,主要用于如视频、高速AD、PCIe、DMA接口等需要高速数据传输的场合。

AXI4:高性能存储映射接口

AXI4接口,它由五个独立的通道构成

1、 读地址
2、 读数据
3、 写地址
4、 写数据
5、 写响应
下面是使用读地址和读数据通道实现读传输过程的示意图:

从图 15.1.1中可以看到,在一个读传输过程中,主机首先在读地址通道给出读地址和控制信号,然后从机由读数据通道返回读出的数据。另外我们需要注意的是,这是一次突发读操作,主机只给出一个地址,从该地址连续突发读出四个数据。

写传输过程如图 15.1.2所示,它用到了写地址、写数据和写响应三个通道。主机在写地址通道给出写地址和控制信号,然后在写数据通道连续突发写四个数据。从机在接收数据之后,在写响应通道给出响应信号。

AXI总线中的每个通道都包含了一组信息信号,还有一个V ALID和一个READY信号。V ALID信号由源端(source)产生,表示当前地址或者数据线上的信息是有效的;而READY信号由目的端(destination)产生,则表示已经准备好接收地址、数据以及控制信息。VALID和READY信号提供了AXI总线中的握手机制,如下图所示:

ACLK为时钟信号,在AXI协议中,所有的输入信号都在是ACLK的上升沿采样,所
有的输出信号必须在ACLK的上升沿之后才能改变。在T1之后,源端将V ALID拉高,表明INFORMA TION信号线上传输的是有效的地址、数据或者控制信息。目的端在T2之后将READY拉高,表明它已经准备好接收数据,此时源端必须保持INFORMA TION数据稳定不变,直到T3时刻进行数据传输。
需要注意的是,源端不允许等目的端的READY信号拉高之后,才将V ALID信号置为有效状态。而且,一旦V ALID拉高,源端必须保持其处于有效状态,直至成功握手(在时钟上升沿检测到V ALID和READY同时为有效状态)。

接下来通过自定义一个AXI4接口的IP核,通过AXI_HP接口对PS端DDR3进行读写测试。

我们在PL内自定义的DDR3 Test IP核作为主设备,通过PS AXI_HP0接口,与DDR控制器进行通信,最终对DDR3存储器进行读写操作。

SDK

1  #include <stdio.h> 
2  #include "xil_cache.h" 
3  #include "xil_printf.h" 
4  #include "xil_io.h" 
5   
6  int main)() 
7  { 
8      int i; 
9      char c; 
10  
11     Xil_DCacheDisable)(); 
12     print("AXI4 PL DDR TEST!\n\r";); 
13  
14     hlwhile(1{){ 
15         scanf("%c"&,&c;); 
16         fif(c==='c'{){ 
17             printf("start\n\r";); 
18             ofor(i=0;i<4096;i=i+4{){ 
19                 printf("%d is %d\n",i(,(int)Xil_In32(0x10000000+i))); 
20             } 
21         } 
22     } 
23  
24     eunreturn 0; 
25 } 

AXI4-Stream协议

AXI4-Stream协议一般被翻译为AXI流协议,是AXI总线的一种演化版本。AXI4流协议作为一个标准接口,用于连接进行数据交换的组件。接口可以用来连接一个单一的主机,主机向接收数据的单一从机发送数据,也可用于连接若干个主机和从机的组件。协议支持共用一组信号线的多个数据流,允许构建一个通用互联。相比于AHB/APB,AXI流协议提出了数据包、数据帧以及传输操作等概念,这也是其被称为流(Stream)的原因。
关于AXI Stream的基本概念解释如下:
传输(Transfer):通过 AXI4 流接口进行的一个单一数据传输。一个单一数据传输由TV ALID和TREADY握手信号定义。
包(Packet):通过 AXI4 流接口被一起传输的一组字节,包类似于 AXI4 的突发。
帧(Frame):一个 AXI4 流中最高级别的字节编组。一帧可以包含很大数量的字节数,例如,一个完整的视频帧缓存。
数据流(Data Stream):从一个源设备到一个目标设备传输的数据。
两个模块之间进行数据传输,需要事先约定好这两个模块之间的传输协议,这是两个信号握手的概念。TV ALID和TREADY信号的握手包含三种情况:TV ALID先于 TREADY 的握手、TREADY先于 TV ALID的握手、TV ALID 和 TREADY 同时发生的握手。

注意该协议中使用上升沿采样
下图中,主机发出了数据和控制信息并将TV ALID 信号置为高。一旦主机驱动了 TV ALID ,主机发出的数据或控制信息必须保持不变,直到从机驱动 TREADY 信号为高表示可以接收数据和控制信息。在这种情况下,一旦从机设置 TREADY 为高,传输就会发生。箭头标示出了传输发生的位置。

下图中,从机在数据和控制信息有效之前驱动TREADY为高。这表示目标设备可以在一个ACLK周期内接收数据和控制信息。在这种情况下,一旦主机驱动 TV ALID 为高,则传输就会发生。箭头标示出了传输发生的位置。

下图中,主机驱动TV ALID为高,从机在同一时钟(ACLK)周期内也驱动TREADY为高。在这种情况下,如图中箭头标注,传输在同一周期内发生。

本次实验我们需要使用Vivado HLS工具生成带有AXI4-Stream接口的IP核,并将此IP核的AXI4-Stream接口连接到“AXI4-Stream to Video Out”模块中的AXI4-Stream接口,如下图所示:

我们重点关注图中的“s_axis_video_tlast”和“s_axis_video_tuser” 信号,其中“s_axis_video_tlast”是AXI4-Stream协议中“TLAST”信号,这个信号设置为高表示一行像素传输结束,“s_axis_video_tuser”是AXI4-Stream协议中的“TUSER”信号,这个信号设置为高表示一帧图像传输开始。时序图如下图所示:

图中的“EOL”表示“End of line”是行传输结束信号,它在一行图像像素传输结束的时候拉高一个时钟周期;图中的“SOF”表示“Start of frame”是帧传输开始信号。它在一帧图像像素传输开始的时候拉高一个时钟周期。

PS和PL的交互方式汇总

最近有个项目关于FPGA加速神经网络,因此需要了解相关PS和pl交互的方法,从而将权重数据存放在PS的DDR中 ,并在需要时完成数据的存入和读出。这是一篇归纳性的文章,具体的还需要自行查阅资料。

https://wallhaven.cc/

PS-PL数据交互方式

1、IO


个数
分布控制
MIO54BANK0, 1PS直接控制
EMIO64BANK2, 3需要PL配置引脚
GPIOAXI-GPIO

MIO :ZYNQ 分为 PS 和 PL 两部分,那么器件的引脚(Pin)资源同样也分成了两部分。ZYNQ PS 中的外设可以通过 MIO(Multiuse I/O,多用输入/输出)模块连接到 PS 端的引脚上,也可以通过 EMIO 连接到 PL 端的引脚。Zynq-7000 系列芯片一般有 54 个 MIO,个别芯片如 7z007s 只有 32 个。

EMIO : PS 和外部设备之间的通信主要是通过复用的输入/输出(Multiplexed Input/Output,MIO)实现的。除此之外,PS 还可以通过扩展的 MIO(Extended MIO,EMIO)来实现与外部设备的连接。EMIO 使用了 PL 的I/O 资源,当 PS 需要扩展超过 54 个引脚的时候可以用 EMIO,也可以用它来连接 PL 中实现的 IP 模块。

在大多数情况下,PS 端经由 EMIO 引出的接口会直接连接到 PL 端的器件引脚上,通过 IO 管脚约束来指定所连接 PL 引脚的位置。通过这种方式,EMIO 可以为 PS 端实现额外的 64 个输入引脚或 64 个带有输出使能的输出引脚。EMIO 还有一种使用方式,就是用于连接 PL 内实现的功能模块(IP 核),此时 PL 端的 IP 作为 PS 端的一个外部设备。(EMIO既可以将ps和pl端的引脚相连,也可以和pl中的模块相连)

PS 与 PL 最主要的连接方式则是一组 AXI 接口。AXI 互联和接口作为 ZYNQ PS 和 PL 之间的桥梁,能够使两者协同工作,进而形成一个完整的、高度集成的系统。

GPIO : AXI GPIO IP 核为 AXI 接口提供了一个通用的输入/输出接口。与 PS 端的 GPIO 不同,AXI GPIO 是一个软核(Soft IP),即 ZYNQ 芯片在出厂时并不存在这样的一个硬件电路,而是由用户通过配置 PL 端的逻辑资源来实现的一个功能模块。而 PS 端的 GPIO 是一个硬核(Hard IP),它是一个生产时在硅片中实现的功能电路。 AXI GPIO IP 模块的左侧实现了一个 32 位的 AXI4-Lite 从接口,用于主机访问 AXI GPIO 内部各通道的寄存器。

AXI GPIO 框图

中断

参考:

https://blog.csdn.net/wangjie36/article/details/116081755

中断是一种当满足要求的突发事件发生时通知处理器进行处理的信号。中断可以由硬件处理单元和外部设备产生,也可以由软件本身产生。对硬件来说,中断信号是一个由某个处理单元产生的异步信号,用来引起处理器的注意。对软件来说,中断还是一种异步事件,用来通知处理器需要改变代码的执行,不过,轮询所产生的中断的过程是同步的。

Zynq 芯片的 PS 部分是基于使用双核 Cortex-A9 处理器和 GIC pl390 中断控制器的 ARM 架构。中断结构与 CPU 紧密链接,并接受来自 I/O 外设(IOP)和可编程逻辑(PL)的中断。

中断控制器架构图

ZYNQ CPU 软件中断(SGI,Software generatedinterrupts):ZYNQ 共有两个 CPU,每个 CPU 具备各自的 16 个软件中断(中断号0-15)(16–26 reserved):被路由到一个或者两个CPU上,通过写ICDSGIR寄存器产生SGI。

CPU私有外设中断(PPI,private peripheralinterrupts ):私有中断是固定的不能修改。这里有 2 个 PL 到 CPU 的快速中断 nFIQ(中断号27-31):每个CPU都有一组PPI,包括全局定时器、私有看门狗定时器、私有定时器和来自PL的FIQ/IRQ。

ZYNQ PS 和 PL 共享中断(SPI,shared peripheralinterrupts):共享中断就是一些端口共用一个中断请求线:(中断号32-95)。由PS和PL上的各种I/O控制器和存储器控制器产生,这些中断信号被路由到相应的CPU, PL部分有16个共享中断,它们的触发方式可以设置。

FIFO

https://zhuanlan.zhihu.com/p/47847664

FIFO类型读接口写接口
AXI Data FIFOAXI4-fullAXI4-full
AXI-Stream FIFOPS axi4-litePL axi-stream
AXI4-Stream Data FIFOaxi-streamaxi-stream

通过AXI-Stream FIFO完成PS和PL部分的数据交互

  • S_AXI, PS读写FIFO数据接口
  • AXI_STR_TXC, 发送控制端口
  • AXI_STR_TXD,发送数据端口
  • AXI_STR_RXD,接收数据端口

读写fifo例程:

写fifo

//write fifo us1
always@(posedge  wrclk, negedge sys_reset_n_i)
begin
    if (!sys_reset_n_i)
    begin
        fifo_wrreq_ddr3_us <= 0 ;
        fifo_data_ddr3_us <= 0 ;
    end
    else
    begin
        if(fifo_prog_full_ddr3_us!= 1) 
            fifo_wrreq_ddr3_us <= 1 ;
        else
            fifo_wrreq_ddr3_us <= 0 ;        
        if(fifo_wrreq_ddr3_us == 1) 
         begin            
             if(fifo_data_ddr3_us < 64'b1111_1111_1111_1111_1111_1111)
                fifo_data_ddr3_us <=fifo_data_ddr3_us + 1 ;                  
             else 
                 fifo_data_ddr3_us <= 0  ; 
         end  
         else
             fifo_data_ddr3_us <= fifo_data_ddr3_us ;                
    end
end
endmodule

读fifo

assign fifo_rdreq_ddr3_ds = !fifo_empty_ddr3_ds;

always@(posedge sys_clk_i,negedge sys_reset_n_i)
begin
    if(!sys_reset_n_i)
         rd_ck_flag_cp<= 1'b0;
    else
    begin
        if (fifo_q_ddr3_ds_r!==fifo_q_ddr3_ds)
            rd_ck_flag_cp<= 1'b1;  
        else
            rd_ck_flag_cp<= 1'b0;
    end    
end
 //jiao yan cuo wu ji shu    jiao yan wei zi zeng       
always@(posedge sys_clk_i,negedge sys_reset_n_i)
begin
    if(!sys_reset_n_i)
    begin
         rd_ck_cnt <= 64'b0;
         fifo_q_ddr3_ds_r <= 64'b0 ;  
    end
    else  
    begin
        if(rd_ck_flag_cp==1)
            rd_ck_cnt <=rd_ck_cnt+1'b1;
        else
            rd_ck_cnt <= rd_ck_cnt;
             
         if ( ( fifo_rdreq_ddr3_ds==1 ) && (fifo_q_ddr3_ds_r < 64'b1111_1111_1111_1111_1111_1111) )
            
             fifo_q_ddr3_ds_r <= fifo_q_ddr3_ds_r + 1'b1 ; 
         else
             fifo_q_ddr3_ds_r <= 64'b0; 
     end
end

endmodule

BRAM

在 ZYNQ SOC 开发过程中,PL 和 PS 之间经常需要做数据交互。对于传输速度要求较高、数据量大、地址连续的场合,可以通过 AXI DMA 来完成。而对于数据量较少、地址不连续、长度不规则的情况,此时 AXIDMA 便不再适用了。针对这种情况,可以通过 BRAM 来进行数据的交互。

BRAM(Block RAM)是 PL 部分的存储器阵列,PS 和 PL 通过对 BRAM 进行读写操作,来实现数据的交互。在 PL 中,通过输出时钟、地址、读写控制等信号来对 BRAM 进行读写操作;而在 PS 中,处理器并不需要直接驱动 BRAM 的端口,而是通过 AXI BRAM 控制器来对 BRAM 进行读写操作。AXI BRAM 控制器是集成在 Vivado 设计软件中的软核,可以配置成 AXI4-lite 接口模式或者 AXI4 接口模式。

AXI4 接口模式的 BRAM 控制器支持的数据位宽为 32 位、64 位、128 位、512 位和 1024 位,而 AXI4-Lite 接口仅支持 32 位数据位宽。由图 14.1.1 可知,PS 通过 AXI4-Lite 接口访问 BRAM,当使能 ECC 选项时,ECC 允许 AXI 主接口检测和纠正 BRAM 块中的单位和双位错误。AXI BRAM 控制器作为 AXI 总线的从接口,和 AXI 主接口实现互联,来对 BRAM 进行读写操作。针对不同的应用场合,该 IP 核支持单次传输和突发传输两种方式。

PS 端的 M_AXI_GP0 作为主端口,与 PL 端的 AXI BRAM 控制器 IP 核和 PL 读 BRAMIP 核(pl_bram_rd)通过 AXI4 总线进行连接。其中,AXI 互联 IP(AXI Interconnect)用于连接 AXI 存储器映射(memory-mapped)的主器件和从器件;AXI BRAM 控制器作为 PS 端读写 BRAM 的 IP 核;PL 读BRAM IP 核是我们自定义的 IP 核,实现了 PL 端从 BRAM 中读出数据的功能,除此之外,PS 端通过 AXI总线来配置该 IP 核读取 BRAM 的起始地址和个数等。

DMA

DMA(Direct Memory Access,直接存储器访问)是计算机科学中的一种内存访问技术。它允许某些计算机内部的硬件子系统可以独立地直接读写系统内存,而不需中央处理器(CPU)介入处理。DMA 是一种快速的数据传送方式,通常用来传送数据量较多的数据块,很多硬件系统会使用 DMA,包括硬盘控制器、绘图显卡、网卡和声卡,在使用高速 AD/DA 时使用 DMA 也是不错的选择。DMA 是用硬件实现存储器与存储器之间或存储器与 I/O 设备之间直接进行高速数据传输。使用 DMA时,CPU 向 DMA 控制器发出一个存储传输请求,这样当 DMA 控制器在传输的时候,CPU 执行其它操作,传输操作完成时 DMA 以中断的方式通知 CPU。

为了发起传输事务,DMA 控制器必须得到以下数据:
• 源地址 — 数据被读出的地址
• 目的地址 — 数据被写入的地址
• 传输长度 — 应被传输的字节数

DMA 存储传输的过程如下:

  1. 为了配置用 DMA 传输数据到存储器,处理器发出一条 DMA 命令
  2. DMA 控制器把数据从外设传输到存储器或从存储器到存储器,而让 CPU 腾出手来做其它操作。
  3. 数据传输完成后,向 CPU 发出一个中断来通知它 DMA 传输可以关闭了。

ZYNQ 提供了两种 DMA,一种是集成在 PS 中的硬核 DMA,另一种是 PL 中使用的软核 AXI DMAIP。在 ARM CPU 设计的过程中,已经考虑到了大量数据搬移的情况,因此在 CPU 中自带了一个 DMA 控制器 DAMC,这个 DAMC 驻留在 PS 内,而且必须通过驻留在内存中的 DMA 指令编程,这些程序往往由CPU 准备,因此需要部分的 CPU 参与。DMAC 支持高达 8 个通道,所以多个 DMA 结构的核可以挂在单个DMAC 上。DAMC 与 PL 的连接是通过 AXI_GP 接口,这个接口最高支持到 32 位宽度,这也限制了这种模式下的传输速率,理论最高速率为 600MB/s。这种模式不占用 PL 资源,但需要对 DMA 指令编程,会增加软件的复杂性。为了获取更高的传输速率,可以以空间换时间,在 PL 中添加 AXI DMAIP 核,并利用 AXI_HP 接口完成高速的数据传输。

为了获取更高的传输速率,可以以空间换时间,在 PL 中添加 AXI DMAIP 核,并利用 AXI_HP 接口完成高速的数据传输,通过 PL 的 DMA 和 AXI_HP 接口的传输适用于大块数据的高性能传输,带宽高。各种接口方式的比较如下表所示:

ZYNQ 开发板上使用 PL 的 AXI DMA IP 核从 DDR3 中读取数据,并将数据写回到 DDR3 中。

在实际应用中,DMA 一般与产生数据或需求数据的 IP 核相连接,该 IP 核可以是带有 Stream 接口的高速的 AD(模拟转数字)或 DA(数字转模拟) IP 核。不失一般性,在本次实验中,我们使用 AXI4 Stream Data FIFO IP 核来充当这类 IP 进行 DMA 环回实验。

AXI Direct Memory Access重要端口说明:

  • S_AXI_LITE: 配置DMA工作模式
  • M_AXI_MM2S:DDR到DMA数据接口
  • M_AXI_S2MM:DMA数据到DDR接口
  • S_AXIS_S2MM: 接收的DMA数据输出端口
  • M_AXIS_MM2S: 想通过DMA输出的数据写入端口

DDR3

通过对AXI HP接口的操作来实现

AXI 的英文全称是 Advanced eXtensible Interface,即高级可扩展接口,它是 ARM 公司所提出的 AMBA(Advanced Microcontroller Bus Architecture)协议的一部分。

AXI4 协议支持以下三种类型的接口:
1、 AXI4:高性能存储映射接口。
2、 AXI4-Lite:简化版的 AXI4 接口,用于较少数据量的存储映射通信。
3、 AXI4-Stream:用于高速数据流传输,非存储映射接口。

AXI4 协议支持突发传输,主要用于处理器访问存储器等需要指定地址的高速数据传输场景。AXI-Lite为外设提供单个数据传输,主要用于访问一些低速外设中的寄存器。而 AXI-Stream 接口则像 FIFO 一样,数据传输时不需要地址,在主从设备之间直接连续读写数据,主要用于如视频、高速 AD、PCIe、DMA 接口等需要高速数据传输的场合。

读传输

AXI 总线中的每个通道都包含了一组信息信号,还有一个 VALID 和一个 READY 信号,VALID 信号由源端(source)产生,表示当前地址或者数据线上的信息是有效的;而 READY 信号由目的端(destination)产生,则表示已经准备好接收地址、数据以及控制信息。

通过自定义一个 AXI4 接口的 IP 核,通过 AXI_HP 接口对 PS 端 DDR3 进行读写测试。

AXI_HP总线:只能单向传输,从PL到PS端,适用于大数据传输。

系统图

PL实现AXI4接口,通过S_AXI_HP接口读取ps侧DDR3数据. 例程功能:PL,PS向指定地址写数据,对方来读.

AXI-DMA:实现从PS内存到PL高速传输高速通道AXI-HP<—->AXI-Stream的转换
AXI-Datamover:实现从PS内存到PL高速传输高速通道AXI-HP<—->AXI-Stream的转换,只不过这次是完全由PL控制的,PS是完全被动的。
AXI-VDMA:实现从PS内存到PL高速传输高速通道AXI-HP<—->AXI-Stream的转换,只不过是专门针对视频、图像等二维数据的。
AXI-CDMA IP: 这个是由PL完成的将数据从内存的一个位置搬移到另一个位置,无需CPU来插手。这个和我们这里用的Stream没有关系

自定义AXI接口IP

一般应用场景在于PS对某些寄存器的配置,传输少量的数据信息。

HLS打包ip报错解决方法:

在HLS进行ip核打包的时候,出现了Vivado fails to export IPs with the error message “Bad lexical cast: source type value could not be interpreted as target”错误 ,在xilinx官网找到了解决办法:

以下是原帖

https://support.xilinx.com/s/question/0D52E00006uxy49SAA/vivado-fails-to-export-ips-with-the-error-message-bad-lexical-cast-source-type-value-could-not-be-interpreted-as-target?language=en_US

https://support.xilinx.com/s/question/0D52E00006uxnnFSAQ/2022-timestamp-overflow-error-2201011128-is-an-invalid-argument-please-specify-an-integer-value?language=en_US

错误是 Vivado 在导出 IP 步骤 (export_design) 中失败。

上面是错误截图,从2021年变成2022年之后就开始出现了。

我尝试重新启动并运行相同的命令来导出以前测试过的设计上的 xo 文件。他们都提示同样的错误

你好,

我正在尝试从 Vitis HLS 导出 Vivado IP(我尝试了 2020.1、2020.2 和 2021.1 版本)。一切都过去正常工作。

但是,现在它会打印以下错误消息:

错误:“2201011128”是无效参数。请指定一个整数

我注意到机器生成的 tcl 脚本使用当前日期作为修订的名称。正如您在下面的屏幕截图中看到的

图片

上面的屏幕截图来自 2021 年 12 月 31 日的设计,效果很好。以下来自 2022 年无效的设计。

图片似乎数字 22 造成了整数溢出。因为 2^31 小于当前修订号。

这个问题有简单的解决方法吗?

请尝试以下解决方法解决此问题:

1.修改vitis_​​hls项目解决目录下的run_ippack.tcl文件

XX\test\solution1\impl\ip\run_ippack.tcl

示例修改: 

 设置修订版“2201012126”-> 设置修订版“2001012126”

但修改后:

修改修订后,如果我再次开始综合,更改将被覆盖。我如何完成生成设计?

最终解决方案:

我将计算机的日期设置回滚到 2021 年,并禁用了自动时间和日期选项。