在Altera SoC DE1板卡上跑完整的卷积神经网(🐵)络
这次为大家详细展示一个利用卷积神经网(🍨)络实现(🙉)图片自动分类的例程(⛷)。
神经网络的优点:自动从数据中学习经验知识,无需复杂的模型和算法。
缺点:有监督学习,需要大量的带标签数据(✔);参数量太少时容易过拟合,泛(⛹)化能力差,参数量太大时(✳)训练收敛很慢(有可(🍖)能需要几个月到几年)。
为了克服上述缺点,人们(💏)发掘了各种计算资源,包括多核CPU、GPU、DSP、ASIC、FPGA,甚至使用模拟电(♿)路。
使用CPU实现卷积神经网络比较方便调试(⛽),但性能太差,一般人们都选用(🧞)更快的GPU实现。目前开源的框架大多都支持GPU,如伯克利大学Caffe和Google Convnet。
微软在2015年2月宣布使(😙)用Stratix V完成了CNN加速器,处理 CIFAR10 图片速度可达每秒2300多张。
这里我们也使用CIFAR10图片数据,在Cyclone V板子上跑一个卷积神经网络CNN demo。由于板子上计算资源太少(DSP Slice只有80多个),实现完整的网络不太现实,只能在FPGA上实现基本计算单元,然后由HPS统一(🆘)调度。性能预期不会太高,后面给出。
CIFAR10图(🙉)片都是什么呢?先来张图!
有兴(💖)趣的朋友可以到官(⛄)网下载(🚂)(CIFAR10官(🐤)网)(⛎)。上面提到过,CNN是有监督学习(🥩)系统,需要大(🗾)量带label的数据,CIFAR10就是这样一(🕰)个开放的数据库,提供了60000张不(🛰)同类别的图片,分为10个类(如(👧)上图左侧所示),每个类别有600张图。这(👸)个数据集不算特别大,适合(🍖)在嵌入式平台上实现。而更大的数据集有ImageNet-1000(ImageNet官网),拥有120多万张高清无码大图,我下载到硬盘,占用了近200GB空间(只能忍痛将其他rmvb和avi删掉了)!
有朋友会问,不用这些数据行不行,我们的智能手机里面照片能(🅱)不能用于CNN做训练?(🙏)
答案是可以的,只是你的数据集很不“均匀”,采样不够“完备”,训练出的模型是真实模型的“有偏估计”,而(🐌)上述两个(🛑)数据集经过了种种考验(🚡),已经是学术界公认的优质数据集,一年一度的ILSVRC比赛就采用了这些数据集。
说完数据,再说模型。先来看一张经典(🎻)的CNN结构:
这是世界上第一个将CNN实用化的(🦕)例子,实现了手写(🏁)体字母自动识别。在这个CNN模型中,可以看到输入是一张32 x 32的二维图像,经过卷(✝)积层(Convolution)、下采样层(Subsampling,也(😸)称Pooling)、全连接层(Full Connection,也称Inner Product)后,得到一组概率密度,我们选其中概率最大的元素作为该模型对输入图像的分类结果。所以实现CNN时,只需要实现三种基本算法:卷积、下(🔳)采(🈲)样、矩阵乘。除此之外,每层(💧)输出都可选择是否经过(🍢)非线性变换,常用的非线性变换有(🌵)ReLU和Sigmoid,前者计算较为简(🍛)单,使用较为广泛。
Caffe框架中提供了专门(📳)为cifar10数(🔔)据定制(🤙)的模型,是(🏒)用proto格式写的,我们的demo也基于这个模型。内容如下:
name: "CIFAR10_quick_test"input: "data"input_dim: 1input_dim: 3input_dim: 32input_dim: 32layers {name: "conv1"type: CONVOLUTIONbottom: "data"top: "conv1"blobs_lr: 1blobs_lr: 2convolution_param {num_output: 32pad: 2kernel_size: 5stride: 1}}layers {name: "pool1"type: POOLINGbottom: "conv1"top: "pool1"pooling_param {pool: MAXkernel_size: 3stride: 2}}layers {name: "relu1"type: RELUbottom: "pool1"top: "pool1"}layers {name: "conv2"type: CONVOLUTIONbottom: "pool1"top: "conv2"blobs_lr: 1blobs_lr: 2convolution_param {num_output: 32pad: 2kernel_size: 5stride: 1}}layers {name: "relu2"type: RELUbottom: "conv2"top: "conv2"}layers {name: "pool2"type: POOLINGbottom: "conv2"top: "pool2"pooling_param {pool: AVEkernel_size: 3stride: 2}}layers {name: "conv3"type: CONVOLUTIONbottom: "pool2"top: "conv3"blobs_lr: 1blobs_lr: 2convolution_param {num_output: 64pad: 2kernel_size: 5stride: 1}}layers {name: "relu3"type: RELUbottom: "conv3"top: "conv3"}layers {name: "pool3"type: POOLINGbottom: "conv3"top: "pool3"pooling_param {pool: AVEkernel_size: 3stride: 2}}layers {name: "ip1"type: INNER_PRODUCTbottom: "pool3"top: "ip1"blobs_lr: 1blobs_lr: 2inner_product_param {num_output: 64}}layers {name: "ip2"type: INNER_PRODUCTbottom: "ip1"top: "ip2"blobs_lr: 1blobs_lr: 2inner_product_param {num_output: 10}}layers {name: "prob"type: SOFTMAXbottom: "ip2"top: "prob"}复制代码
可见,上述模型经过(🈷)了3个卷积层(conv1, conv2, conv3),每个卷积层后面都跟着下采样层(pool1, pool2, pool3),之后有两个全连接(😹)层(ip1, ip2),最后一层prob为SOFTMAX分类层,是计算概率(🔢)密度的,这里我们不需要关心。
下面三张图分别统计了CNN模型(🕦)各层的参数量、数据量和计算量。
可(🌃)以看(😡)出,卷积层的参数量很少,但数据量很大;全连接层刚好相反,参数量较大,但数据量很少。
通过计算量统计发现conv2计算量最大,其次(🌈)是conv3和conv1。全连接层的(💿)计算量相对卷积层较小,但不可忽略。其他层(pool1, pool2以及各级relu)由于计算量太小,本设计中没有将其实现为Open CL kernel,而是直接CPU端实现。
综上所述,我们重点实现(😸)两个算法:卷积和(🌀)矩阵乘,分别对应卷积层、全连接层的实现。
在DE1-SOC上(📷)我利用了友晶提供的Open CL BSP,支持C语言开发FPGA。
卷积层计算kernel函数如下:
__attribute__((num_compute_units(4)))__kernelvoid conv(__global float * a, __global float * b, __global float * c, const int M, const int N, const int K){int gx = get_global_id(0);int gy = get_global_id(1);float tmp=0.0f;for(int x = 0; x < K; x ++){for(int y = 0; y < K; y ++){tmp += a[(gx + x) * M + (gy + y)] * b[x * K + y];}}复制代码
全连接层计算采用矩阵乘(🎒)实现,kernel函数如下:
__attribute__((num_compute_units(4)))__kernelvoid gemm(__global float * a, __global float * b, __global float * c, const int M, const int N, const int K){int gx = get_global_id(0);int gy = get_global_id(1);int sy = get_global_size(1);int sx = get_global_size(0);int s = sx * sy;for(int x = gx; x < M; x += sx){for(int y = gy; y < N; y += sy){float tmp=0.0f;for(int z = 0; z < K; z++){tmp += a[z * M + x] * b[y * K + z];}c[y * M + x] = tmp;}}}复制代码
编译kernel函数需要使用Altera SDK for OpenCL,我用的(🔇)版本是14.0.0.200,申请了两个月的license。编译使用命令行aoc,得到*.aocx文件。
Open CL编译输出报告中给出了资源占(💌)用情况:
+--------------------------------------------------------------------+; Estimated Resource Usage Summary ;+----------------------------------------+---------------------------+; Resource + Usage ;+----------------------------------------+---------------------------+; Logic utilization ; 83% ;; Dedicated logic registers ; 46% ;; Memory blocks ; 57% ;; DSP blocks ; 25% ;+----------------------------------------+---------------------------;复制代(🦃)码
可见,逻辑资源、存储器资源消耗较为明显,而DSP资源(🤫)并未用尽,说明还有优化的空间。
编译主程序需要使用SoCEDS,我用的版本为14.0.2.274,也是命令行方式,在工程目录下执行make,结束(📬)后得到可执行文件cnn。
将这两个文件拷贝到SD卡,按(🌛)照前面的博客对板(🔪)子进行设置,将CNN的模型、CIFAR10数(👿)据也拷贝到SD卡中,板子上电,mount SD卡到/mnt,执行cnn,得到输出如下:
<div class="blockcode"><blockquote>Please input the number of images(1~100):100Loading data...OK!Constructing CNN...OK!Begin calculation...Elapsed Time = 141.861 s.Real Label = 3(cat), Calc Label = 3(cat), error count = 0Real Label = 8(ship), Calc Label = 8(ship), error count = 0Real Label = 8(ship), Calc Label = 8(ship), error count = 0Real Label = 0(airplane), Calc Label = 0(airplane), error count = 0Real Label = 6(frog), Calc Label = 6(frog), error count = 0Real Label = 6(frog), Calc Label = 6(frog), error count = 0Real Label = 1(automobile), Calc Label = 1(automobile), error count = 0Real Label = 6(frog), Calc Label = 6(frog), error count = 0Real Label = 3(cat), Calc Label = 3(cat), error count = 0Real Label = 1(automobile), Calc Label = 1(automobile), error count = 0Real Label = 0(airplane), Calc Label = 0(airplane), error count = 0Real Label = 9(truck), Calc Label = 9(truck), error count = 0Real Label = 5(dog), Calc Label = 5(dog), error count = 0Real Label = 7(horse), Calc Label = 7(horse), error count = 0Real Label = 9(truck), Calc Label = 9(truck), error count = 0Real Label = 8(ship), Calc Label = 8(ship), error count = 0Real Label = 5(dog), Calc Label = 5(dog), error count = 0Real Label = 7(horse), Calc Label = 7(horse), error count = 0Real Label = 8(ship), Calc Label = 8(ship), error count = 0Real Label = 6(frog), Calc Label = 6(frog), error count = 0Real Label = 7(horse), Calc Label = 7(horse), error count = 0Real Label = 0(airplane), Calc Label = 2(bird), error count = 1Real Label = 4(deer), Calc Label = 4(deer), error count = 1Real Label = 9(truck), Calc Label = 9(truck), error count = 1Real Label = 5(dog), Calc Label = 4(deer), error count = 2Real Label = 2(bird), Calc Label = 3(cat), error count = 3Real Label = 4(deer), Calc Label = 4(deer), error count = 3Real Label = 0(airplane), Calc Label = 0(airplane), error count = 3Real Label = 9(truck), Calc Label = 9(truck), error count = 3Real Label = 6(frog), Calc Label = 6(frog), error count = 3Real Label = 6(frog), Calc Label = 6(frog), error count = 3Real Label = 5(dog), Calc Label = 5(dog), error count = 3Real Label = 4(deer), Calc Label = 4(deer), error count = 3Real Label = 5(dog), Calc Label = 5(dog), error count = 3Real Label = 9(truck), Calc Label = 9(truck), error count = 3Real Label = 2(bird), Calc Label = 3(cat), error count = 4Real Label = 4(deer), Calc Label = 7(horse), error count = 5Real Label = 1(automobile), Calc Label = 9(truck), error count = 6Real Label = 9(truck), Calc Label = 9(truck), error count = 6Real Label = 5(dog), Calc Label = 5(dog), error count = 6Real Label = 4(deer), Calc Label = 4(deer), error count = 6Real Label = 6(frog), Calc Label = 6(frog), error count = 6Real Label = 5(dog), Calc Label = 5(dog), error count = 6Real Label = 6(frog), Calc Label = 6(frog), error count = 6Real Label = 0(airplane), Calc Label = 0(airplane), error count = 6Real Label = 9(truck), Calc Label = 9(truck), error count = 6Real Label = 3(cat), Calc Label = 5(dog), error count = 7Real Label = 9(truck), Calc Label = 9(truck), error count = 7Real Label = 7(horse), Calc Label = 7(horse), error count = 7Real Label = 6(frog), Calc Label = 6(frog), error count = 7Real Label = 9(truck), Calc Label = 9(truck), error count = 7Real Label = 8(ship), Calc Label = 8(ship), error count = 7Real Label = 0(airplane), Calc Label = 2(bird), error count = 8Real Label = 3(cat), Calc Label = 3(cat), error count = 8Real Label = 8(ship), Calc Label = 8(ship), error count = 8Real Label = 8(ship), Calc Label = 8(ship), error count = 8Real Label = 7(horse), Calc Label = 7(horse), error count = 8Real Label = 7(horse), Calc Label = 7(horse), error count = 8Real Label = 4(deer), Calc Label = 3(cat), error count = 9Real Label = 6(frog), Calc Label = 3(cat), error count = 10Real Label = 7(horse), Calc Label = 7(horse), error count = 10Real Label = 3(cat), Calc Label = 5(dog), error count = 11Real Label = 6(frog), Calc Label = 6(frog), error count = 11Real Label = 3(cat), Calc Label = 3(cat), error count = 11Real Label = 6(frog), Calc Label = 6(frog), error count = 11Real Label = 2(bird), Calc Label = 2(bird), error count = 11Real Label = 1(automobile), Calc Label = 1(automobile), error count = 11Real Label = 2(bird), Calc Label = 2(bird), error count = 11Real Label = 3(cat), Calc Label = 3(cat), error count = 11Real Label = 7(horse), Calc Label = 9(truck), error count = 12Real Label = 2(bird), Calc Label = 2(bird), error count = 12Real Label = 6(frog), Calc Label = 6(frog), error count = 12Real Label = 8(ship), Calc Label = 8(ship), error count = 12Real Label = 8(ship), Calc Label = 8(ship), error count = 12Real Label = 0(airplane), Calc Label = 0(airplane), error count = 12Real Label = 2(bird), Calc Label = 2(bird), error count = 12Real Label = 9(truck), Calc Label = 0(airplane), error count = 13Real Label = 3(cat), Calc Label = 3(cat), error count = 13Real Label = 3(cat), Calc Label = 2(bird), error count = 14Real Label = 8(ship), Calc Label = 8(ship), error count = 14Real Label = 8(ship), Calc Label = 8(ship), error count = 14Real Label = 1(automobile), Calc Label = 1(automobile), error count = 14Real Label = 1(automobile), Calc Label = 1(automobile), error count = 14Real Label = 7(horse), Calc Label = 7(horse), error count = 14Real Label = 2(bird), Calc Label = 2(bird), error count = 14Real Label = 5(dog), Calc Label = 7(horse), error count = 15Real Label = 2(bird), Calc Label = 2(bird), error count = 15Real Label = 7(horse), Calc Label = 7(horse), error count = 15Real Label = 8(ship), Calc Label = 8(ship), error count = 15Real Label = 9(truck), Calc Label = 9(truck), error count = 15Real Label = 0(airplane), Calc Label = 0(airplane), error count = 15Real Label = 3(cat), Calc Label = 4(deer), error count = 16Real Label = 8(ship), Calc Label = 8(ship), error count = 16Real Label = 6(frog), Calc Label = 6(frog), error count = 16Real Label = 4(deer), Calc Label = 4(deer), error count = 16Real Label = 6(frog), Calc Label = 6(frog), error count = 16Real Label = 6(frog), Calc Label = 6(frog), error count = 16Real Label = 0(airplane), Calc Label = 2(bird), error count = 17Real Label = 0(airplane), Calc Label = 0(airplane), error count = 17Real Label = 7(horse), Calc Label = 7(horse), error count = 17Classify Score = 83 %.上(🤜)面的执行流程(🥖)是这(💩)样的,首先输入(🙎)测试样本数(🍤)目(1到100),由于DE1板子(😃)FPGA端SDRAM容量较小,难以加载全部测试数据(10000张图片),故每次最多装入(🦍)100张图片。之后载入(🀄)数据到HPS内存(🔬),然后开始构建CNN模型,构建(🍋)过程中也实现了Open CL的初始化。构建完毕,将输入图(🎺)像依次通过CNN,得到一系列分类结果,与标签进行对比,统计错误分类个数,计算分类准确率。
经过测试,分类准确率达到83%,与Caffe测试结果一致。
经(🏾)过以上测试,可以得到结论:
(1)使用Open CL可以很方便地移植高级语言编写的算法;
(2)CNN在移植过程中需要考虑实际硬件,定制合适的模型和数据;
(3)Cyclone 5逻辑资源较少(85K,Open CL kernel占用了83%),如果希望进一步提高计算速度,一方面可以选用高性能器件(如Stratix V、Arria 10),另一方面可以使用RTL自己搭建计算系统。
以上图文内容(🏡)均是EEWORLD论坛网友zhaoyongke原创,在此(🏼)感谢。
欢迎微博@EEWORLD
如果你也写过此类原创干(👑)货请关注微(🔥)信订(👄)阅号(ID:(😈)eeworldbbs,将你的原创发至:bbs_service@eeworld.com.cn,一经入选,我(➗)们将帮你登上头条!
与更多(📰)行业内网友进行交流(😱)请登陆EEWORLD论坛。
【免费看很黄a片试看120秒的相关新闻】