雷达信号处理脉冲压缩算法GPU实现及加速（含完整代码）

下面展示GPU端完成雷达信号处理脉冲压缩算法加速。
为探索GPU部署雷达信号处理提供参考。

脉冲压缩算法 在GPU实现，模拟LFM线性调频信号，完成GPU端 cuda加速
最终与matlab答案进行正确性验证对比

#include <stdio.h>
#include <stdlib.h>
#include <math.h>
#include <cufft.h>
#include"cuda_runtime_api.h"
#include"device_launch_parameters.h"#define M_PI 3.14159265358979323846
/*
#define BATCH 1
#define SIZE 2048
*/#define BATCH 32
#define SIZE 1080 #include "sys/time.h"
double what_time_is_it_now() {struct timeval time;if (gettimeofday(&time, NULL)) {return 0;}return (double)time.tv_sec + (double)time.tv_usec * .000001;
}__global__ void complexMulKernel(const cuComplex* r_sf, const cuComplex* r_hf, cuComplex* sot1, int num) {int i = blockIdx.x * blockDim.x + threadIdx.x;if (i < num) {sot1[i] = cuCmulf(r_sf[i], r_hf[i]);}
}int main() {double start1, end1,start2, end2,start3, end3;double t1 = 10e-6;double b = 25e6;double k = b / t1;double fs = 200e6;double ts = 1 / fs;double n = t1 / ts;// int n = 2048;cufftComplex* st;cufftComplex* ht;cufftComplex* sf;cufftComplex* hf;cufftComplex* sot1;cufftComplex* sot;cudaMallocHost((void**)&st, SIZE * BATCH * sizeof(cufftComplex));cudaMallocHost((void**)&ht, SIZE * BATCH * sizeof(cufftComplex));cudaMallocHost((void**)&sf, SIZE * BATCH * sizeof(cufftComplex));cudaMallocHost((void**)&hf, SIZE * BATCH * sizeof(cufftComplex));cudaMallocHost((void**)&sot1, SIZE * BATCH * sizeof(cufftComplex));cudaMallocHost((void**)&sot, SIZE * BATCH * sizeof(cufftComplex));//QueryPerformanceCounter(&start_time);cufftComplex* d_st;cufftComplex* d_ht;cufftComplex* d_sf;cufftComplex* d_hf;cufftComplex* d_st2;cufftComplex* d_ht2;cufftComplex* d_sf2;cufftComplex* d_hf2;cufftComplex* d_sot1;cufftComplex* d_sot;cufftComplex* d_sot1_out;cudaMalloc((void**)&d_st, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_ht, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_sf, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_st2, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_ht2, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_sf2, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_hf, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_sot1, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_sot, SIZE * BATCH * sizeof(cufftComplex));cudaMalloc((void**)&d_sot1_out, SIZE * BATCH * sizeof(cufftComplex));// generate linear frequency-modulated signal stdouble t_min = -t1 / 2;double t_max = t1 / 2;double t_step = (t_max - t_min) / (n - 1);double t = t_min;for (int i = 0; i < SIZE ; i++){if (i >= 2000) {st[i].x = 0;st[i].y = 0;}else {st[i].x = cos(M_PI * k * t * t);st[i].y = sin(M_PI * k * t * t);t += t_step;}}//输出st//for (int i = 0; i < SIZE; i++)//{//    printf("(%f, %f)\n", st[i].x, st[i].y);//}// generate matched filter htt = t_min;for (int i = 0; i < SIZE ; i++){if (i >= 2000) {ht[i].x = 0;ht[i].y = 0;}else {ht[i].x = cos(-M_PI * k * t * t);ht[i].y = sin(-M_PI * k * t * t);t += t_step;}}cudaMemcpy(d_st, st, SIZE * BATCH * sizeof(cufftComplex), cudaMemcpyHostToDevice);cudaMemcpy(d_st2, st, SIZE * BATCH * sizeof(cufftComplex), cudaMemcpyHostToDevice);cudaMemcpy(d_ht, ht, SIZE * BATCH * sizeof(cufftComplex), cudaMemcpyHostToDevice);// printf("%f+%f", ht[20481].x, ht[20481].y);// Create cuFFT planscufftHandle plan_st, plan_ht, plan_sot;cufftPlan1d(&plan_st, SIZE , CUFFT_C2C, BATCH);cufftPlan1d(&plan_ht, SIZE , CUFFT_C2C, BATCH);cufftPlan1d(&plan_sot, SIZE , CUFFT_C2C, BATCH);start1 = what_time_is_it_now();// Perform forward FFT on st and htfor(int i=0;i<1000;i++){cufftExecC2C(plan_st, d_st, d_sf, CUFFT_FORWARD);//  cufftExecC2C(plan_st, d_st2, d_sf2, CUFFT_FORWARD);}end1 = what_time_is_it_now();printf(" fft time : %f   ms\n ", 1000 * (end1 - start1) /1000);cufftExecC2C(plan_ht, d_ht, d_hf, CUFFT_FORWARD);cufftComplex* r_sf= (cufftComplex*)malloc(SIZE * BATCH * sizeof(cufftComplex));cufftComplex* r_hf = (cufftComplex*)malloc(SIZE * BATCH * sizeof(cufftComplex));cudaMemcpy(r_sf, d_sf, SIZE * BATCH * sizeof(cufftComplex), cudaMemcpyDeviceToHost);cudaMemcpy(r_hf, d_hf, SIZE * BATCH * sizeof(cufftComplex), cudaMemcpyDeviceToHost);//输出d_st/d_hf/*  for (int i = 0; i < SIZE; i++){printf("(%f, %f)\n", r_hf[i].x, r_hf[i].y);}*/// Multiply frequency domain signalsdim3 threadsPerBlock(256);dim3 numBlocks((2048 + threadsPerBlock.x - 1) / threadsPerBlock.x);start3 = what_time_is_it_now();for(int i=0;i<1000;i++){complexMulKernel << <numBlocks, threadsPerBlock >> > (d_sf, d_hf, d_sot1, SIZE * BATCH);}end3 = what_time_is_it_now();printf(".* time : %f   ms\n ", 1000 * (end3 - start3) /1000);//输出sot1/*for (int i = 0; i < SIZE; i++){printf("(%f, %f)\n", sot1[i].x, sot1[i].y);}*/// Perform inverse FFT on sot1start2 = what_time_is_it_now();for(int i=0;i<1000;i++){cufftExecC2C(plan_sot, d_sot1, d_sot1_out, CUFFT_INVERSE);}end2 = what_time_is_it_now();printf("ifft time : %f   ms\n ", 1000 * (end2 - start2)/1000 );cufftComplex* r_sot = (cufftComplex*)malloc(SIZE * sizeof(cufftComplex));cudaMemcpy(r_sot, d_sot1_out, SIZE * sizeof(cufftComplex), cudaMemcpyDeviceToHost);//输出sot1//for (int i = 0; i < SIZE; i++)//{//    printf("(%f, %f)\n", r_sot[i].x/2048, r_sot[i].y/2048);  //matlabd的 ifft 是把结果除以信号长度得出来的  (逆fft)!!!!!!//}// Shift the resultint half = 2048 / 2;cufftComplex* temp = (cufftComplex*)malloc(half * sizeof(cufftComplex));// Copy the first half of the arrayfor (int i = 0; i < half; i++){temp[i] = r_sot[i];r_sot[i] = r_sot[i + half];r_sot[i + half] = temp[i];}// Output the results// printf("r_sot values:\n");for (int i = 0; i < SIZE; i++){//      printf("(%f, %f)\n", r_sot[i].x/2048, r_sot[i].y/2048);}printf("GPU运行时间：  %f ms\n", 1000 * (end1 - start1)/1000+1000 * (end2 - start2)/1000+1000 * (end3 - start3)/1000);//   ------------------验证误差 < 0.05% -----------------------FILE* fp;cufftComplex* Mix;int i;// 动态分配内存Mix = (cufftComplex*)malloc(SIZE * sizeof(cufftComplex));if (Mix == NULL) {printf("Failed to allocate memory.\n");return 1;}fp = fopen("/home/xtic/MTS/Radar/pc.txt", "r"); // 打开文件a.txt，只读模式if (fp == NULL) {printf("Failed to open file a.txt.\n");return 1;}// 循环读取文件中的复数值for (i = 0; i < SIZE; i++) {float real, imag;fscanf(fp, "%f %fi", &real, &imag);Mix[i].x = real;Mix[i].y = imag;}// 关闭文件fclose(fp);// 测试输出for (i = 0; i < SIZE; i++) {// if (r_sot[i].x/2048 - Mix[i].x > 0.0005) {if (r_sot[i].x/2048 - Mix[i].x > 0.0005) {//            printf(" false!!!!!");} }printf("\nsuccess! 误差小于0.05%\n");/*for (i = 0; i < SIZE; i++) {printf("(%f)\n", Mix[i].x);}printf("\n");*///  释放内存free(Mix);// Clean upcufftDestroy(plan_st);cufftDestroy(plan_ht);cufftDestroy(plan_sot);// free(st); free(ht);// free(sf);// free(hf);// free(sot1);// free(sot);// free(temp);return 0;
}

雷达信号处理脉冲压缩算法GPU实现及加速（含完整代码）相关推荐

【模型压缩】深度卷积网络的剪枝和加速（含完整代码）
作者 | 贝壳er 研究 | 数据挖掘与异常检测出品 | AI蜗牛车 " 记录一下去年12月份实验室的一个工作:模型的剪枝压缩,虽然模型是基于yolov3的魔改,但是剪枝的对象还是CBL层 ...
【雷达信号处理】脉冲多普勒PD及其MATLAB实现
这是目录 1 原理介绍 1.1 脉冲多普勒过程 1.2 信号模型 1.3 PD的实现 1.4 相参和非相参累积 2 实验内容 2.1 参数 3 MATLAB实现参考文献 1 原理介绍 1.1 脉冲多 ...
区块链参考资源, 雷达信号处理
区块链参考资源 https://mp.weixin.qq.com/s/bkLBaTs3pSQyYg9YgRnG8w 盘点近期区块链媒体.论坛.公号资源 https://mp.weixin.qq.com ...
matlab中的方波信号图片_基于Matlab的雷达信号处理仿真
这是一个较为基础的图文教程(含仿真代码):利用MATLAB设计经典的雷达数字信号处理.该系统具备对雷达目标回波的处理能力,能够从噪声中将目标检测出来,并提取目标的距离.速度.角度信息.教程分五节完成, ...
mimo雷达信号处理_雷达学术入门脉冲雷达信号处理概述
Reviewed by :@甜草莓 @Robert Zhou: 前置知识:概率论与统计学.面向人群:本科生.研究生/信号处理博士. 编者:对于信号处理来说,雷达和通信一直是一体两面,从MIMO通信到M ...
雷达信号处理-雷达应用
<雷达信号处理基础>Mark A. Rechards 读书笔记 ---------------------- 信号处理的分类:信号调节和干扰抑制,成像,检测和后处理. 雷达的时间尺度快时 ...
【雷达仿真 | FMCW TDMA-MIMO毫米波雷达信号处理仿真（可修改为DDMA-MIMO）】
本文编辑:调皮哥的小助理本文引用了CSDN雷达博主@XXXiaojie的文章源码(https://blog.csdn.net/Xiao_Jie1),加以修改和注释,全面地.详细地阐述了FMCW TD ...
读书笔记 | 自动驾驶中的雷达信号处理（第9章汽车雷达的应用概述）
本文编辑:调皮哥的小助理大家好,我是调皮哥,又和大家见面了,时间过得很快,到目前为止,本次读书笔记的内容是最后一篇了,相信大家通过之前文章的阅读,已经对自动驾驶中的雷达信号处理.雷达数据处理.人工智 ...
雷达信号处理算法：静态杂波滤除（附MATLAB代码和数据）
本文编辑:调皮哥的小助理本期文章将介绍三种雷达信号处理常用的静态杂波滤方法的基本原理,分别是零速通道置零法.动目标显示(MTI)以及相量均值相消算法(平均相消算法),并分析了静态杂波的滤除效果,以及 ...

雷达信号处理脉冲压缩算法GPU实现及加速（含完整代码）

雷达信号处理脉冲压缩算法GPU实现及加速（含完整代码）相关推荐

最新文章

热门文章