sunwen/Aurora
1
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

Fix fftshift and ifftshift memory bug.

Browse Source
This commit is contained in:
kradchen
2023-12-08 17:32:00 +08:00
parent 0229a132b2
commit 4eeb6c4988
2 changed files with 25 additions and 3 deletions
Download Patch File Download Diff File Expand all files Collapse all files

4
src/Function2D.cu
View File

@@ -1194,7 +1194,7 @@ void Aurora::fftshift(CudaMatrix &aMatrix){
} else { } else {
int copySize = aMatrix.getDimSize(0) / 2 + 1; int copySize = aMatrix.getDimSize(0) / 2 + 1;
float *data = nullptr; float *data = nullptr;
cudaMalloc((void **)&data, copySize * aMatrix.getValueType()); cudaMalloc((void **)&data,copySize*sizeof(float)*aMatrix.getValueType());
memcpyColKernel<<<aMatrix.getDimSize(1), 256>>>( memcpyColKernel<<<aMatrix.getDimSize(1), 256>>>(
aMatrix.getData(), data, copySize * aMatrix.getValueType(), aMatrix.getData(), data, copySize * aMatrix.getValueType(),
aMatrix.getDimSize(0) * aMatrix.getValueType(), aMatrix.getDimSize(0) * aMatrix.getValueType(),
@@ -1220,7 +1220,7 @@ void Aurora::ifftshift(CudaMatrix &aMatrix){
} else { } else {
int copySize = aMatrix.getDimSize(0) / 2 + 1; int copySize = aMatrix.getDimSize(0) / 2 + 1;
float *data = nullptr; float *data = nullptr;
cudaMalloc((void **)&data, copySize * aMatrix.getValueType()); cudaMalloc((void **)&data,copySize*sizeof(float)*aMatrix.getValueType());
memcpyColKernel<<<aMatrix.getDimSize(1), 256>>>( memcpyColKernel<<<aMatrix.getDimSize(1), 256>>>(
aMatrix.getData()+(copySize-1)* aMatrix.getValueType(), aMatrix.getData()+(copySize-1)* aMatrix.getValueType(),
data, copySize * aMatrix.getValueType(), data, copySize * aMatrix.getValueType(),

24
test/Function2D_Cuda_Test.cpp
View File

@@ -684,6 +684,28 @@ TEST_F(Function2D_Cuda_Test, fft)
} }
} }
{
float * data = new float[36];
for (size_t i = 0; i < 18; i++)
{
data[i*2] = i%9+1;
data[i*2+1] = 1;
}
auto fm = Aurora::Matrix::fromRawData(data,9,2,1,Aurora::Complex);
auto fmD = fm.toDeviceMatrix();
Aurora::fftshift(fm);
Aurora::fftshift(fmD);
for (size_t i = 0; i < fm.getDataSize()*2; i++)
{
EXPECT_FLOAT_AE(fm[i], fmD.getValue(i));
}
Aurora::ifftshift(fm);
Aurora::ifftshift(fmD);
for (size_t i = 0; i < fm.getDataSize()*2; i++)
{
EXPECT_FLOAT_AE(fm[i], fmD.getValue(i));
}
}
{ {
float * data = new float[250000]; float * data = new float[250000];
for (size_t i = 0; i < 250000; i++) for (size_t i = 0; i < 250000; i++)
@@ -721,7 +743,7 @@ TEST_F(Function2D_Cuda_Test, fft)
} }
Aurora::ifftshift(fm); Aurora::ifftshift(fm);
Aurora::ifftshift(fmD); Aurora::ifftshift(fmD);
for (size_t i = 0; i < fm.getDataSize(); i++) for (size_t i = 0; i < fm.getDataSize()*2; i++)
{ {
EXPECT_FLOAT_AE(fm[i], fmD.getValue(i)); EXPECT_FLOAT_AE(fm[i], fmD.getValue(i));
} }