百度 AI Infra (2)
Q: OpenMP并行区怎么开?
使用 #pragma omp parallel 编译器指令开启并行区域,编译器在遇到该指令时会 fork 出一组线程来并行执行花括号内的代码块。
常见用法:
#pragma omp parallel for:自动将 for 循环的迭代空间均匀分配到各线程,是最常用的并行化方式#pragma omp parallel sections:将不同代码段分配到不同线程并行执行num_threads(N):手动指定线程数,默认使用环境变量OMP_NUM_THREADS或系统核心数
调度策略(通过 schedule 子句控制循环分配方式):
static:预先均分迭代,适合每次迭代计算量相同的场景dynamic:运行时按需分配,适合迭代计算量不均的场景,但有调度开销guided:动态分配但块大小递减,折中方案
注意事项:并行区内要注意变量的共享/私有属性(shared/private/firstprivate/lastprivate),错误的默认共享可能导致数据竞争。编译时需加 -fopenmp(GCC)或 /openmp(MSVC)。
Q: for(int i=0;i<100;i++){sum++;} 在并行区内和串行区的运行结果是否相同?如何变成相同的?
不相同。多线程并行执行时 sum++ 存在数据竞争(race condition)。sum++ 实际包含”读-改-写”三个步骤,多个线程可能同时读到旧值再各自加 1 写回,导致更新丢失。最终结果不确定,且通常小于 100。
解决方法(按推荐优先级排列):
reduction 子句(最优方案):
#pragma omp parallel for reduction(+:sum)。每个线程维护私有副本,循环结束后自动合并。无额外同步开销,编译器自动优化。原子操作:
#pragma omp atomic修饰sum++。硬件级原子指令保证读-改-写不可分割。开销比 reduction 大(每次迭代都有原子操作),但适合不规则的更新模式。私有副本手动合并:每个线程用
private变量累加,循环结束后在 critical 区合并。代码稍复杂但语义清晰。互斥锁:
#pragma omp critical保护sum++。这是最慢的方案,因为所有线程在此串行化,并行度几乎为零。仅适合临界区内有复杂操作的情况。
性能对比:reduction > atomic >> critical。对于简单累加操作,reduction 几乎是零额外开销。
Q: GPU中的内存架构是什么?
GPU 内存架构是分层设计,核心原则是”越靠近计算单元的存储越快但越小”。以 NVIDIA A100 为例:
| 层级 | 容量 | 延迟 | 带宽 | 作用域 |
|---|---|---|---|---|
| 寄存器 | 每SM 256KB | 1 cycle | - | 线程私有 |
| 共享内存/L1 | 每SM 192KB(可配) | ~20 cycles | ~19 TB/s | Block内共享 |
| L2 Cache | 40MB | ~200 cycles | ~5 TB/s | 全GPU共享 |
| HBM(全局内存) | 80GB | ~400 cycles | 2 TB/s | 全GPU共享 |
各层详细说明:
- 寄存器(Register File):每个 CUDA Core 私有,速度最快。每个 SM 有 64K 个 32-bit 寄存器。寄存器用完会溢出到 Local Memory(实际在 HBM 中),性能急剧下降。
- 共享内存/L1 Cache:SM 内所有线程共享的片上存储,程序员可显式管理(共享内存部分)。L1 和共享内存在 Ampere 架构共享 192KB 物理空间,可配置比例。
- L2 Cache:所有 SM 共享,硬件自动管理。A100 有 40MB L2,对全局内存访问起缓存作用。
- 全局内存(HBM):容量最大延迟最高。HBM2e 提供约 2TB/s 带宽,但延迟高达 400-800 周期。
- 常量内存和纹理内存:只读,有专用缓存路径,适合广播式读取(所有线程读同一地址零开销)。
优化核心思路:尽可能让数据在片上存储(寄存器 > 共享内存)中复用,减少对 HBM 的访问。
Q: Cache映射的分类和特点?
Cache 映射决定了主存块在 Cache 中的放置规则,影响命中率和硬件复杂度。
1. 直接映射(Direct Mapped):
- 规则:主存块只能映射到 Cache 中的唯一固定行(block_addr % cache_lines)
- 优点:硬件最简单,只需一次比较即可判断命中,查找延迟最低
- 缺点:冲突率高(thrashing),两个映射到同一行的频繁访问块会反复驱逐对方
- 典型应用:L1 指令 Cache(访问模式相对固定)
2. 全相联映射(Fully Associative):
- 规则:主存块可放置在 Cache 的任意行
- 优点:冲突率最低,空间利用率最高
- 缺点:需要与所有行并行比较(CAM 内容寻址),硬件复杂度和功耗高
- 典型应用:TLB(条目数少,通常 32-64 项)
3. 组相联映射(Set Associative):
- 规则:Cache 分为若干组(set),主存块映射到特定组,组内任意行可放置
- N 路组相联:每组有 N 行。如 8-way 表示每组 8 行
- 兼顾冲突率和硬件复杂度,是现代 CPU Cache 的主流方案
- 典型配置:L1 Cache 4-8 路,L2 Cache 8-16 路,L3 Cache 12-16 路
替换策略:当组内满时需要淘汰一行,常用 LRU(最近最少使用)、Pseudo-LRU、Random。
Q: C++智能指针的分类和用法?
C++11 引入智能指针替代裸指针的手动内存管理,通过 RAII 机制在析构时自动释放资源。
shared_ptr:共享所有权,内部维护引用计数(存储在控制块 control block 中)。每次拷贝 +1,每次析构 -1,计数归零时释放资源。注意:(1) 引用计数的增减是原子操作,因此 shared_ptr 本身的拷贝/析构是线程安全的,但其指向的对象的读写不是;(2) 控制块额外占用约 16-32 字节;(3) 优先使用 make_shared 一次分配对象和控制块,减少内存分配次数。
unique_ptr:独占所有权,不可拷贝(只可 std::move 转移)。无控制块、无引用计数,与裸指针开销相同(零开销抽象)。适合明确单一所有者的场景,是默认首选的智能指针。支持自定义删除器(如释放 FILE*、CUDA memory 等)。
weak_ptr:弱引用,不增加引用计数。必须从 shared_ptr 构造,使用时通过 lock() 提升为 shared_ptr。主要用途:(1) 打破 shared_ptr 循环引用(如双向链表、观察者模式);(2) 缓存场景中检查对象是否仍存活。
选择策略:默认用 unique_ptr -> 需要共享所有权时用 shared_ptr -> 需要观察但不拥有时用 weak_ptr。
Q: C++多态是什么?怎么实现的?
多态是指同一接口在不同对象上表现出不同行为,是面向对象编程的核心特性之一。
编译时多态(静态多态):
- 函数重载:同名函数不同参数列表,编译期根据实参类型选择调用版本
- 模板:泛型编程,编译期实例化具体类型。CRTP(奇异递归模板模式)可实现”静态虚函数”效果,无运行时开销
运行时多态(动态多态):
- 基类声明
virtual函数,派生类override重写 - 实现机制:编译器为每个含虚函数的类生成虚函数表(vtable),表中存放该类各虚函数的实际地址。每个对象开头有一个 vptr(通常 8 字节)指向所属类的 vtable
- 调用过程:通过基类指针/引用调用虚函数时 -> 读取对象的 vptr -> 在 vtable 中找到函数地址 -> 间接跳转调用
- 性能开销:一次额外的指针间接寻址(通常 L1 命中,开销约 1-2 ns),在循环热路径中可能影响性能
注意事项:虚函数不能内联(因为地址运行时才确定);虚析构函数是基类的必须(否则 delete 基类指针不会调用派生类析构)。
Q: 动态链接和动态绑定的区别?
两者虽然都有”动态”,但属于完全不同的概念层面:
动态绑定(Dynamic Binding / Late Binding):
- 属于语言层面的概念,是 C++ 运行时多态的核心机制
- 编译时不确定调用哪个函数,运行时根据对象的实际类型(通过 vptr 查 vtable)确定调用目标
- 只对 virtual 函数生效,非虚函数在编译期就已绑定(静态绑定/早绑定)
动态链接(Dynamic Linking):
- 属于操作系统/链接器层面的概念
- 程序运行时才加载共享库(Linux 的
.so、Windows 的.dll、macOS 的.dylib) - 优点:节省磁盘和内存(多个程序共享同一份库代码)、便于库更新(无需重新编译程序)
- 对比静态链接:静态链接在编译时将库代码复制到可执行文件中,体积大但无运行时依赖
- 通过 PLT/GOT(Linux)或 IAT(Windows)实现延迟绑定,首次调用时解析符号地址
Q: C++ STL中map的底层实现?
map 底层是红黑树(Red-Black Tree),一种自平衡二叉搜索树。
核心性质:
- 每个节点是红色或黑色
- 根节点为黑色
- 红色节点的子节点必须为黑色(无连续红色)
- 从根到任意叶子的路径上黑色节点数相同(黑高相同)
性能特征:插入、删除、查找时间复杂度均为 O(log n),元素按 key 有序排列(中序遍历得到升序序列)。
与 unordered_map 对比:map 有序但查找慢(O(log n) vs O(1)),适合需要范围查询或有序遍历的场景;unordered_map 无序但查找快,适合纯粹的键值查找。
内存开销:每个节点除了存 key-value 外还需存储颜色位、左右子节点和父节点指针(约 3 个指针 + 1 bit),比 vector 等连续存储的内存局部性差。
Q: AVL树和红黑树有什么区别?
| 对比维度 | AVL树 | 红黑树 |
|---|---|---|
| 平衡条件 | 严格:左右子树高度差<=1 | 宽松:最长路径<=2*最短路径 |
| 树高 | ~1.44*log(n),更矮 | ~2*log(n),略高 |
| 查找效率 | 略优(树更矮) | 略逊 |
| 插入旋转 | 最多2次旋转 | 最多2次旋转 |
| 删除旋转 | 可能O(log n)次旋转 | 最多3次旋转 |
| 适用场景 | 查找密集型(如数据库索引) | 插入/删除频繁(如STL map/set) |
STL 选择红黑树的原因:实际应用中插入/删除操作频繁,红黑树删除时旋转次数有上界(最多 3 次),而 AVL 树删除可能需要从删除点回溯到根的路径上多次旋转,维护成本更高。在工程实践中,红黑树的综合性能更优。
Q: unordered_map的底层实现?
底层是哈希表(Hash Table),STL 实现使用拉链法(separate chaining)处理冲突。
结构:维护一个 bucket 数组(vector),每个 bucket 是一个链表(或在 C++23 后可能是其他容器)。通过哈希函数将 key 映射到 bucket 索引,冲突元素挂在同一 bucket 的链表上。
性能:
- 平均:插入/删除/查找 O(1)
- 最坏:所有元素哈希到同一个 bucket,退化为 O(n)
- 扩容:当 load_factor(元素数/桶数)超过阈值(默认 1.0)时 rehash,将桶数翻倍并重新分配所有元素,单次 O(n)
注意事项:
- 自定义类型做 key 需要提供 hash 函数和 == 运算符
- 内存不连续(链表节点),cache 局部性比 vector 差
- 元素无序,不支持范围查询
Q: 哈希冲突的解决方法?
哈希冲突不可避免(鸽巢原理),关键在于如何高效处理。
1. 链地址法(Separate Chaining):
- 每个桶维护一个链表(或红黑树),冲突元素追加到链表
- 优点:实现简单,对 load factor 容忍度高(可>1),删除方便
- 缺点:链表节点内存不连续,cache 不友好
- STL unordered_map、Java HashMap(链表长度>8 转红黑树)采用此方案
2. 开放定址法(Open Addressing):
- 冲突时按探测序列在表内寻找下一个空位
- 线性探测:步长 1,简单但易聚集(primary clustering)
- 二次探测:步长为 i^2,减少聚集但不保证找到空位
- 双重哈希:用第二个哈希函数计算步长,分布最均匀
- 优点:数据连续存储 cache 友好,适合小元素
- 缺点:load factor 不能太高(通常<0.75),删除复杂(需要 tombstone 标记)
- Google dense_hash_map、Python dict 采用此方案
3. 其他方法:再哈希法(使用备选哈希函数)、建立公共溢出区(冲突元素统一存放)。
Q: 计算机的存储层次架构?
从快到慢、从小到大的层次结构,利用程序的时间局部性和空间局部性原理。
| 层级 | 典型容量 | 典型延迟 | 带宽 |
|---|---|---|---|
| 寄存器 | ~KB | <1 ns (1 cycle) | - |
| L1 Cache | 32-64 KB | ~1 ns (3-4 cycles) | ~1 TB/s |
| L2 Cache | 256 KB-1 MB | ~3-5 ns (10-12 cycles) | ~500 GB/s |
| L3 Cache | 8-64 MB | ~10-20 ns (30-40 cycles) | ~200 GB/s |
| 主存(DRAM) | 16-256 GB | ~50-100 ns | ~50 GB/s |
| SSD | TB级 | ~100 us | ~5 GB/s |
| HDD | TB级 | ~10 ms | ~200 MB/s |
设计哲学:每一层作为下一层的缓存。程序 90% 的时间访问 10% 的数据(局部性原理),因此小而快的存储可以有效服务大部分访问。Cache miss penalty 是性能优化的核心考量。
Q: CPU中应用程序员可见的寄存器分类?PC寄存器的作用?
程序员可见寄存器分类(以 x86-64 为例):
- 通用寄存器(GPR):RAX/RBX/RCX/RDX/RSI/RDI/R8-R15 等 16 个 64 位寄存器,用于数据运算和地址计算。某些寄存器有约定用途(如 RAX 存函数返回值,RSP 为栈指针)
- 段寄存器:CS/DS/SS/ES/FS/GS,64 位模式下大部分已不直接使用(FS/GS 用于 TLS)
- 标志寄存器(RFLAGS):存放算术运算结果状态(ZF 零标志、CF 进位、OF 溢出等)和控制位(IF 中断使能)
- SIMD 寄存器:XMM0-15(128bit)/ YMM0-15(256bit)/ ZMM0-31(512bit),用于向量化计算
程序计数器(PC / RIP):存放下一条将要执行的指令在内存中的地址。CPU 取指阶段根据 PC 值从内存加载指令,执行后 PC 自动递增(顺序执行)或被跳转指令修改(分支/调用)。在 x86 中称为 RIP(Instruction Pointer),用户态程序不能直接写入,只能通过 jmp/call/ret 等指令间接修改。
Q: 在main函数前运行一个函数怎么实现?
C++ 程序的启动实际上在 main 之前有一个初始化阶段(由 CRT startup code 执行),可以利用这个机制:
全局对象的构造函数:定义全局对象,其构造函数在
main之前执行。原理:CRT 初始化代码遍历.init_array段中的构造函数指针并逐一调用。执行顺序:同一翻译单元内按定义顺序,不同翻译单元间顺序未定义(Static Initialization Order Fiasco)。GCC constructor 属性:
__attribute__((constructor)) void func() {...}。可指定优先级__attribute__((constructor(priority))),数字越小越先执行(101-65535 用户可用)。底层原理同样是在.init_array段注册函数指针。静态变量初始化:全局/静态变量的初始化表达式在
main前求值。但注意 C++ 中非 constexpr 的静态初始化顺序跨翻译单元未定义。
注意事项:main 前的代码不应依赖其他翻译单元的全局对象状态,否则可能触发 SIOF 问题。解决方案:使用函数内局部静态变量(Construct On First Use Idiom)。
Q: 如何设置条件断点?
GDB 中:break 行号 if 条件,例如 break 10 if i==50。也可以先设置普通断点再添加条件:condition 断点编号 表达式。
高级用法:
- 条件可以是任意 C 表达式:
break func if strcmp(name, "target")==0 - 忽略前 N 次命中:
ignore 断点编号 N - 命中时自动执行命令:
commands 断点编号后输入命令序列
IDE 中:右键断点 -> 编辑条件 -> 输入布尔表达式。VS Code、CLion 等均支持。
性能注意:条件断点会显著降低调试速度(每次到达断点都要评估条件表达式)。对于高频执行的循环,考虑使用 watchpoint(watch 变量)监控变量变化,或使用硬件断点。
Q: 怎么进行堆栈监视?
GDB 调试:
bt(backtrace):查看完整调用栈,包含函数名、参数值、源文件行号bt full:显示每个栈帧的局部变量值frame N:切换到第 N 层栈帧查看其上下文info locals:当前栈帧的局部变量info args:当前函数的参数值up/down:在栈帧间移动
运行时检测工具:
- AddressSanitizer(ASan):编译时加
-fsanitize=address,检测栈缓冲区溢出、use-after-return、栈上越界。运行时开销约 2x。 - Valgrind/Memcheck:无需重新编译,检测内存泄漏、未初始化读取、越界访问。开销约 10-20x。
- Stack canary:编译器自动插入(
-fstack-protector),在返回地址前放置随机值,函数返回时检查是否被覆盖。
Q: 程序在内存中的分段?
从高地址到低地址的典型布局(Linux x86-64):
1 | 高地址 ┌─────────────────────┐ |
关键点:栈和堆相向增长,中间有大量未映射空间(由 ASLR 随机化基址)。mmap 区域通常在栈和堆之间,用于动态库加载和大块 malloc。
Q: STL中deque是怎么实现的?
deque(双端队列)采用分段连续存储架构:
内部结构:
- 一个中控数组(map):存放指针,每个指针指向一块固定大小的连续缓冲区(通常 512 字节或 sizeof(T)<512 时为 512/sizeof(T) 个元素)
- 每个缓冲区存放若干个连续的元素
设计优势:
- 逻辑上连续,物理上分段,支持 O(1) 的头尾插入/删除(只需在头/尾缓冲区操作)
- 不像 vector 需要整体 realloc,扩容时只需新增缓冲区并更新中控数组
- 随机访问 O(1):通过计算目标元素所在缓冲区的索引和偏移实现
与 vector 对比:
- vector 头部插入 O(n)(需移动全部元素),deque O(1)
- vector 连续内存 cache 友好,deque 跨缓冲区时有间接寻址开销
- vector 适合随机访问密集,deque 适合头尾操作频繁的场景(如 BFS 队列)
Q: 寄存器和Cache哪个更快?Cache和主存的实现方式有什么区别?
寄存器更快:访问延迟为 0-1 个时钟周期(在流水线中可视为即时),因为寄存器文件直接与 ALU 相连,无需经过缓存层。
Cache vs 主存的硬件实现差异:
| 特性 | Cache (SRAM) | 主存 (DRAM) |
|---|---|---|
| 存储单元 | 6 个晶体管/bit(双稳态触发器) | 1 个晶体管 + 1 个电容/bit |
| 刷新 | 不需要(状态稳定) | 需要每 64ms 刷新一次(电容漏电) |
| 速度 | 快(1-20 ns) | 慢(50-100 ns) |
| 密度 | 低(面积大) | 高(面积小) |
| 成本 | 高(约 $/GB 是 DRAM 的 100 倍) | 低 |
| 功耗 | 待机功耗高(漏电流) | 刷新有功耗但待机低 |
相同点:都是易失性存储(断电数据丢失),都通过电信号读写。
为什么要分层:SRAM 太贵不能做大容量,DRAM 太慢不能直接喂 CPU。Cache 作为 CPU 和主存之间的缓冲,利用局部性原理以少量 SRAM 覆盖大部分访问。
Q: CUDA编程模型中的内存分类?
CUDA 编程模型中提供了多层次的内存抽象,对应 GPU 硬件的不同存储层级:
| 内存类型 | 作用域 | 生命周期 | 物理位置 | 延迟 |
|---|---|---|---|---|
| 寄存器 | 线程私有 | 线程 | 片上 | 1 cycle |
| 本地内存 | 线程私有 | 线程 | DRAM(HBM) | ~400 cycles |
| 共享内存 | Block内共享 | Block | 片上(SM) | ~20 cycles |
| 全局内存 | 所有线程 | 应用 | DRAM(HBM) | ~400 cycles |
| 常量内存 | 所有线程(只读) | 应用 | DRAM+缓存 | 1-400 cycles |
| 纹理内存 | 所有线程(只读) | 应用 | DRAM+缓存 | 1-400 cycles |
关键注意点:
- 本地内存名称有误导性,实际位于 HBM 中(与全局内存一样慢),仅在寄存器溢出或动态索引数组时使用
- 共享内存是优化的关键工具:手动管理的 L1 缓存,可用于线程间数据共享和全局内存访问合并
- 常量内存总量 64KB,全部缓存时延迟等同 L1。适合所有线程读同一值的广播场景
- 纹理内存有 2D 空间局部性优化,适合图像处理中的邻域访问模式
Q: CUDA中__global__关键字的作用是什么?
__global__ 是 CUDA 函数类型限定符之一,修饰的函数称为核函数(Kernel)。
核心特性:
- 由 Host(CPU)发起调用,在 Device(GPU)上由大量线程并行执行
- 调用语法:
kernel<<<gridDim, blockDim, sharedMemBytes, stream>>>(args...) - 返回类型必须为
void(异步执行,无法直接返回值) - 不支持递归(Compute Capability < 5.0)、函数指针受限、不支持可变参数
三种函数类型限定符对比:
__global__:Host 调用,Device 执行(kernel)__device__:Device 调用,Device 执行(设备函数,可被 kernel 调用)__host__:Host 调用,Host 执行(普通 CPU 函数,默认)- 可组合使用
__host__ __device__生成 CPU 和 GPU 两个版本
注意事项:kernel 是异步执行的,调用后 CPU 立即返回继续执行后续代码。需要 cudaDeviceSynchronize() 或 stream 同步来等待 kernel 完成。
Q: 如何获取GPU允许的最大线程数量?
使用 cudaGetDeviceProperties(&prop, device_id) 获取设备属性结构体,其中包含:
1 | cudaDeviceProp prop; |
实际限制:maxThreadsPerBlock=1024 只是硬件上限。实际能用多少线程还取决于 kernel 的寄存器使用量和共享内存使用量(影响 occupancy)。可用 CUDA Occupancy Calculator 或 __launch_bounds__ 指导编译器优化。
Q: 贪心算法和动态规划的区别?
两者都通过子问题求解原问题,但策略和适用条件不同:
| 对比维度 | 贪心算法 | 动态规划 |
|---|---|---|
| 决策方式 | 每步取当前局部最优 | 考虑所有子问题组合取全局最优 |
| 回溯 | 不回溯,一旦选择不可撤销 | 通过备忘录/表格保留所有可能 |
| 最优性 | 不保证全局最优(需证明贪心选择性质) | 保证全局最优 |
| 时间复杂度 | 通常 O(n) 或 O(n log n) | 通常 O(n^2) 或更高 |
| 空间复杂度 | O(1) | O(n) 或 O(n^2)(存表) |
| 适用条件 | 贪心选择性质 + 最优子结构 | 最优子结构 + 重叠子问题 |
典型贪心问题:区间调度、Huffman 编码、Dijkstra(非负权边)、最小生成树(Kruskal/Prim)
典型 DP 问题:最长公共子序列、背包问题、编辑距离、最短路径(含负权 Bellman-Ford)
判断方法:如果问题能证明”当前最优选择不会影响未来最优子结构”,用贪心;否则需要 DP 枚举所有可能。
Q: 图搜索算法有哪些?
基础遍历算法:
BFS(广度优先搜索):使用队列逐层扩展。性质:最先找到的路径是最短路径(无权图)。时间 O(V+E),空间 O(V)。应用:最短路径、层序遍历、连通性检测。
DFS(深度优先搜索):使用栈或递归深入到底再回溯。时间 O(V+E),空间 O(V)。应用:拓扑排序、强连通分量(Tarjan)、环检测、回溯搜索。
最短路径算法:
- Dijkstra:单源最短路,贪心策略(优先队列),不支持负权边。时间 O((V+E)logV)。
- Bellman-Ford:单源最短路,支持负权边,可检测负环。时间 O(VE)。
- Floyd-Warshall:全源最短路,DP 思路。时间 O(V^3)。
- **A***:启发式搜索,f(n) = g(n) + h(n)。当启发函数可容许时保证最优。适合有明确目标的图搜索(如路径规划)。
选择依据:无权图用 BFS;正权图用 Dijkstra;含负权用 Bellman-Ford;全对最短路用 Floyd;有启发信息用 A*。