diff --git a/src/Function2D.cu b/src/Function2D.cu index 4cc3e95..8a4fa8f 100644 --- a/src/Function2D.cu +++ b/src/Function2D.cu @@ -1335,9 +1335,14 @@ CudaMatrix Aurora::fft(const CudaMatrix &aMatrix, long aFFTSize){ size_t needCopySize = (ColEleCount>>(aMatrix.getData(), data, needCopySize, aMatrix.getDimSize(0),ColEleCount); + cudaMalloc((void**)&data, sizeof(float)*2*bufferSize); + if (aMatrix.isComplex()){ + complexCopyKernel<<>>(aMatrix.getData(), data, needCopySize, aMatrix.getDimSize(0),ColEleCount); + } + else{ + complexFillKernel<<>>(aMatrix.getData(), data, needCopySize, aMatrix.getDimSize(0),ColEleCount); + } auto ret = Aurora::CudaMatrix::fromRawData(data,ColEleCount,aMatrix.getDimSize(1),1,Complex); ExecFFT(ret,0); return ret; diff --git a/test/Function2D_Cuda_Test.cpp b/test/Function2D_Cuda_Test.cpp index a68c455..584345d 100644 --- a/test/Function2D_Cuda_Test.cpp +++ b/test/Function2D_Cuda_Test.cpp @@ -640,12 +640,6 @@ TEST_F(Function2D_Cuda_Test, fft) EXPECT_FLOAT_EQ(fourDecimalRound(ifftResult[3].real()),2.0); EXPECT_FLOAT_EQ(fourDecimalRound(ifftResult[11].real()),1.0); EXPECT_FLOAT_EQ(fourDecimalRound(ifftResult[13].real()),2.0); - // Aurora::fftshift(fftrett); - // EXPECT_FLOAT_EQ(0.0729, fourDecimalRound(result[6].real())); - // EXPECT_FLOAT_EQ(2.4899, fourDecimalRound(result[7].imag())); - // EXPECT_FLOAT_EQ(-2., fourDecimalRound(result[10].real())); - // EXPECT_FLOAT_EQ(-0.2245, fourDecimalRound(result[11].imag())); - auto fftrett2 = Aurora::fft(ma,5); auto fftrett2H =fftrett2.toHostMatrix(); @@ -668,7 +662,31 @@ TEST_F(Function2D_Cuda_Test, fft) } { float *input = new float[20]{1,1,0,2,2,0,1,1,0,2,1,1,0,2,2,0,1,1,0,2}; - auto ma = Aurora::Matrix::fromRawData(input,10,2); + auto ma = Aurora::Matrix::fromRawData(input,10,1, 1,Aurora::Complex); + auto maD = ma.toDeviceMatrix(); + auto ret1 = Aurora::fft(ma); + auto ret2 = Aurora::fft(maD); + for (size_t i = 0; i < ret1.getDataSize(); i++) + { + EXPECT_FLOAT_AE(ret1[i], ret2.getValue(i)); + } + ret1 = Aurora::fft(ma,12); + ret2 = Aurora::fft(maD,12); + for (size_t i = 0; i < ret1.getDataSize(); i++) + { + EXPECT_FLOAT_AE(ret1[i], ret2.getValue(i)); + } + ret1 = Aurora::fft(ma,8); + ret2 = Aurora::fft(maD,8); + for (size_t i = 0; i < ret1.getDataSize(); i++) + { + EXPECT_FLOAT_AE(ret1[i], ret2.getValue(i)); + } + + } + { + float *input = new float[20]{1,1,0,2,2,0,1,1,0,2,1,1,0,2,2,0,1,1,0,2}; + auto ma = Aurora::Matrix::fromRawData(input,10,1,1, Aurora::Complex); auto maD = ma.toDeviceMatrix(); auto ret1 = Aurora::ifft(Aurora::fft(ma),12); auto ret2 = Aurora::ifft(Aurora::fft(maD),12);