一个高性能的LDPC(低密度奇偶校验)置信传播解码器的CUDA实现。
本项目实现了一个基于CUDA的LDPC置信传播解码器,用于解码LDPC编码的数据。通过GPU并行计算,实现了高效的解码性能。
- 高性能CUDA实现:利用GPU并行计算加速解码过程
- 模块化设计:清晰的代码结构,易于维护和扩展
- RAII内存管理:自动管理CUDA设备内存,避免内存泄漏
- 完善的错误处理:全面的CUDA错误检查和异常处理
- 可配置性:支持JSON配置文件和命令行参数
- 跨平台构建:支持CMake构建系统
BPDecoder_3-1/
├── CMakeLists.txt # CMake构建配置
├── README.md # 项目文档
├── include/ # 公共头文件
│ ├── LDPCParameters.h # LDPC参数类
│ ├── CUDABPDecoder.h # CUDA解码器类
│ ├── MemoryManager.h # CUDA内存管理
│ ├── CSVReader.h # CSV文件读取
│ └── ErrorHandler.h # 错误处理
├── src/ # 源代码
│ ├── main.cpp # 主程序
│ ├── LDPCParameters.cpp
│ ├── CUDABPDecoder.cpp
│ ├── MemoryManager.cpp
│ ├── CSVReader.cpp
│ └── ErrorHandler.cpp
├── cuda/ # CUDA内核代码
│ ├── BPKernels.cu # BP解码内核实现
│ └── BPKernels.h # 内核声明
├── tests/ # 测试代码
│ └── test_ldpc.cpp
└── data/ # 测试数据
├── offsetWeight.csv # 偏移和权重数据
├── columnIndexMap.csv # 列索引映射
└── llr.csv # LLR值
- CUDA Toolkit: 11.0 或更高版本
- C++编译器: 支持C++14标准
- CMake: 3.18 或更高版本
- GPU: NVIDIA GPU,计算能力 6.0+
# 创建构建目录 mkdir build cd build # 配置项目 cmake .. # 构建 cmake --build . --config Release # 运行 ./bin/Release/BPDecoder_CUDA
项目保留了原有的Visual Studio项目文件(BPDecoder_3.sln),可以直接使用Visual Studio打开并构建。
Usage: BPDecoder_CUDA [options]
Options:
-h, --help 显示帮助信息
-c, --config <file> 配置文件路径 (JSON格式)
-d, --data <dir> 数据目录路径 (默认: data/)
-i, --iterations <n> 最大迭代次数 (默认: 50)
-t, --threads <n> 每块线程数 (默认: 256)
--device <id> CUDA设备ID (默认: 0)
-v, --verbose 详细输出
# 使用默认参数 ./BPDecoder_CUDA # 指定数据目录和迭代次数 ./BPDecoder_CUDA -d data/ -i 50 # 使用配置文件 ./BPDecoder_CUDA -c config.json # 详细输出模式 ./BPDecoder_CUDA -v -i 100
支持JSON格式的配置文件:
{
"ldpc_parameters": {
"block_length": 16200,
"parity_length": 2880,
"max_iterations": 50,
"num_edges": 49319,
"num_threads": 256,
"device_id": 0
}
}项目需要三个CSV文件作为输入:
- offsetWeight.csv: 包含行/列偏移和权重信息
- columnIndexMap.csv: 列索引映射数据
- llr.csv: 对数似然比(LLR)值
所有CSV文件使用逗号分隔,每行一个数据项或多个逗号分隔的值。
封装LDPC解码所需的所有参数和数据:
bp_decoder::LDPCParameters params;
params.loadDataFiles("data/");核心解码器类,管理整个解码流程:
bp_decoder::CUDABPDecoder decoder(params);
bp_decoder::DecodeResult result = decoder.decode();RAII模式的CUDA内存管理:
bp_decoder::CUDABuffer<float> buffer(1000); // 自动分配和释放 buffer.copyFromHost(hostData);
统一的错误处理机制:
// CUDA错误检查宏 CHECK_CUDA_ERROR(cudaMalloc(&ptr, size)); CHECK_KERNEL_LAUNCH();
本项目实现了标准的LDPC置信传播解码算法,包含三个核心步骤:
- 校验节点更新 (Lr): 计算从校验节点到变量节点的消息
- 变量节点更新 (LQ): 计算后验对数似然比
- 消息更新 (Lq): 计算从变量节点到校验节点的消息
- 每个校验节点/变量节点的更新在独立的CUDA线程中执行
- 使用CUDA事件进行精确的性能计时
- 支持多种GPU架构(SM 6.0 - 8.6)
- 使用
__restrict__关键字优化内存访问 - 合理的线程块大小配置
- 最小化主机与设备之间的数据传输
项目实现了完整的异常层次结构:
BPDecoderException: 基础异常类CUDAException: CUDA相关错误FileException: 文件操作错误ConfigException: 配置相关错误
运行测试:
cd build
ctest本项目仅供学习和研究使用。
- v3.1: 重构版本
- 模块化代码结构
- RAII内存管理
- 完善的错误处理
- CMake构建支持
- 命令行参数支持
如有问题或建议,请提交Issue或Pull Request。