3. Intel GPU架构:集成显卡与独立显卡、EU执行单元与渲染切片、GPU计算模型
好,咱们今天聊聊Intel的GPU。说实话,很多做嵌入式的朋友一听到GPU,第一反应就是“那是做图形渲染的,跟我搞控制、搞信号处理有什么关系?”
嗯,我以前也这么想。直到我在一个边缘计算项目里,被CPU的算力瓶颈卡得死死的。后来试着把一部分矩阵运算扔到集显上跑,效果出乎意料。从那以后,我再也不敢小看这颗“附赠”的芯片了。
3.1 集成显卡与独立显卡:同根不同命
Intel的GPU分两条线:集成显卡和独立显卡。说白了,集成显卡是跟CPU焊在同一块芯片上的,共享内存。独立显卡则是单独一张卡,有自己的显存和供电。
在嵌入式领域,咱们打交道最多的其实是集成显卡。比如我常用的那些低功耗平台,Tiger Lake、Alder Lake,甚至更早的Apollo Lake,里面都塞了一个Gen系列核显。你想想看,一个SoC搞定CPU+GPU,省面积、省功耗、还省BOM成本,这对嵌入式设计来说太香了。
独立显卡呢?Intel最近几年才开始认真做,比如Arc系列。性能确实猛,但功耗和体积摆在那里。我个人建议,除非你的应用需要高分辨率实时渲染或者大规模并行计算,否则别轻易上独显。嵌入式讲究的是“够用就好”。
核心区别一句话总结:
- 集成显卡:共享系统内存,延迟稍高,带宽受限,但零额外成本。
- 独立显卡:专用显存,高带宽低延迟,但贵、大、费电。
避坑指南: 我曾经在一个工业视觉项目里,为了省成本选了集成显卡。结果发现图像预处理时,CPU和GPU抢内存带宽,导致系统卡顿。后来我强制把GPU的显存占用限制在256MB以内,才稳住。记住,集显不是免费的午餐,你得学会跟它“分蛋糕”。
3.2 EU执行单元与渲染切片:GPU的“肌肉”和“骨架”
讲完了宏观分类,咱们深入看看Intel GPU的内部结构。这里有两个核心概念:EU和Slice。
EU(Execution Unit,执行单元),你可以把它想象成GPU里的一个“小工人”。每个EU内部有多个线程槽位,可以同时处理多个线程。Intel的EU设计很有意思,它不像NVIDIA那样搞大核心,而是用大量小EU堆算力。
我举个例子。在Gen12架构里,一个EU包含8个ALU(算术逻辑单元),每个ALU可以执行SIMD(单指令多数据)操作。这意味着一个EU在一个时钟周期内,可以同时处理8个浮点运算。你想想看,如果芯片里有96个EU,那并行能力有多恐怖?
渲染切片(Render Slice),则是把这些EU组织起来的“骨架”。一个Slice通常包含一组EU、一个L3缓存、以及一些固定功能单元(比如光栅化、深度测试)。多个Slice之间通过一个环形总线连接。
为什么会这样设计?说白了,是为了灵活扩展。低端芯片可以只放一个Slice,高端芯片放四个Slice。我在做功耗优化时,就经常通过控制Slice的开关来调节性能。比如在待机状态下,只保留一个Slice工作,其他全部休眠。
| 架构代际 | 每Slice EU数 | 典型配置 | 我的使用感受 |
|---|---|---|---|
| Gen9 (Skylake) | 24 | 1-3 Slice | 够用,但跑复杂shader会吃力 |
| Gen11 (Ice Lake) | 32 | 1-2 Slice | 能效比提升明显,适合移动端 |
| Gen12 (Tiger Lake) | 48 | 1-2 Slice | 性能翻倍,我开始用它做轻量级AI推理 |
注意: 不是EU越多越好。我曾经在一个项目中,把EU从48个加到96个,结果性能只提升了30%。为什么?因为内存带宽成了瓶颈。EU再多,数据喂不进去也是白搭。所以做系统设计时,一定要算好“算力-带宽”的匹配关系。
3.3 GPU计算模型:从图形到通用计算的跨越
好了,硬件结构讲完了。但咱们搞嵌入式的,更关心的是怎么用它。这就引出了GPU计算模型。
传统的GPU是为图形设计的。图形渲染的特点是:数据量大、计算简单、并行度高。比如你要渲染一帧画面,需要对几百万个像素做同样的变换操作。GPU就是为这种“单指令多数据”模式量身定做的。
后来大家发现,很多非图形任务也符合这个模式。比如矩阵乘法、卷积运算、FFT(快速傅里叶变换)。于是就有了GPGPU(通用GPU计算)。
Intel的GPU计算模型,核心是SIMD(单指令多数据)和SIMT(单指令多线程)的结合。什么意思呢?
- SIMD:一条指令,操作多个数据。比如一条加法指令,同时加8个浮点数。
- SIMT:多个线程,执行同一条指令,但每个线程处理不同的数据。
Intel的EU内部,实际上是把SIMD和SIMT混合使用的。每个EU可以同时运行多个硬件线程,每个线程内部又使用SIMD指令。这种设计的好处是,既能利用线程级并行,又能利用数据级并行。
我举个实际例子。我在做图像去噪时,需要把一张1920x1080的图片分成很多个小块,每个小块交给一个线程去处理。每个线程内部,又用SIMD指令同时处理8个像素。这样一来,整个处理过程就快了很多。
// 伪代码示例:使用Intel GPU进行图像卷积
// 每个线程处理一个像素块
kernel void convolution_kernel(
global float* input,
global float* output,
constant float* filter,
int width, int height)
{
int x = get_global_id(0);
int y = get_global_id(1);
// 使用SIMD指令同时处理多个像素
float8 sum = 0.0f;
for (int i = -1; i <= 1; i++) {
for (int j = -1; j <= 1; j++) {
float8 pixel = vload8(0, input + (y+i)*width + (x+j));
sum += pixel * filter[(i+1)*3 + (j+1)];
}
}
vstore8(sum, 0, output + y*width + x);
}
我的经验: 刚开始写GPU程序时,我总想着把任务拆得越细越好。后来发现不对。每个线程处理的任务太小,调度开销反而成了大头。我现在的习惯是,让每个线程处理至少64个元素,这样EU的利用率才能上去。你想想看,如果每个线程只处理1个元素,那EU大部分时间都在等调度,算力全浪费了。
最后,我想强调一点。Intel的GPU计算模型,跟NVIDIA的CUDA模型有相似之处,但也有区别。最大的区别在于,Intel的GPU更强调向量化。你写代码时,要时刻想着“怎么把数据打包成向量”。如果你习惯了标量编程,刚开始可能会有点不适应。但一旦掌握了这个思路,你会发现性能提升是立竿见影的。
好了,这一章的内容就到这里。下一章咱们聊聊Intel的媒体引擎和编解码器,那又是另一个有意思的话题。