diff --git a/psrdada_cpp/effelsberg/edd/GatedSpectrometer.cuh b/psrdada_cpp/effelsberg/edd/GatedSpectrometer.cuh index 3870f0fbd62297e13e0f1ee7d1fe4ffab45c2bbe..f3502675ea7a0c2b4697430dc7e1419c1d809e2c 100644 --- a/psrdada_cpp/effelsberg/edd/GatedSpectrometer.cuh +++ b/psrdada_cpp/effelsberg/edd/GatedSpectrometer.cuh @@ -269,6 +269,37 @@ struct GatedFullStokesOutput: public OutputDataStream +struct GatedSpectrometerInputParameters + { + /** + * @brief GatedSpectrometerInputParameters defining behavior of GatedSpectrometer + * + * @param buffer_bytes A RawBytes object wrapping a DADA header buffer + * @param nSideChannels Number of side channel items in the data stream, + * @param selectedSideChannel Side channel item used for gating + * @param selectedBit bit of side channel item used for gating + * @param speadHeapSize Size of the spead heap block. + * @param fftLength Size of the FFT + * @param naccumulate Number of samples to integrate in the individual + * FFT bins + * @param nbits Bit depth of the sampled signal + * @param handler Output handler + * + */ + + effelsberg::edd::DadaBufferLayout dadaBufferLayout; + size_t selectedSideChannel; + size_t speadHeapSize; + size_t nSideChannels; + size_t selectedBit; + size_t fft_length; + size_t naccumulate; + unsigned int nbits; + // size_t disable_gate; + }; + + + /** @class GatedSpectrometer @@ -280,25 +311,12 @@ template <class HandlerType, class OutputType > class GatedSpectrometer { public: + /** * @brief Constructor - * - * @param buffer_bytes A RawBytes object wrapping a DADA header buffer - * @param nSideChannels Number of side channel items in the data stream, - * @param selectedSideChannel Side channel item used for gating - * @param selectedBit bit of side channel item used for gating - * @param speadHeapSize Size of the spead heap block. - * @param fftLength Size of the FFT - * @param naccumulate Number of samples to integrate in the individual - * FFT bins - * @param nbits Bit depth of the sampled signal - * @param handler Output handler - * + */ - GatedSpectrometer(const DadaBufferLayout &bufferLayout, - std::size_t selectedSideChannel, std::size_t selectedBit, - std::size_t fft_length, - std::size_t naccumulate, std::size_t nbits, + GatedSpectrometer(const GatedSpectrometerInputParameters &ip, HandlerType &handler); ~GatedSpectrometer(); @@ -346,6 +364,7 @@ private: thrust::device_vector<uint64_cu> &_noOfBitSetsIn_G1); DadaBufferLayout _dadaBufferLayout; + std::size_t _fft_length; std::size_t _naccumulate; std::size_t _selectedSideChannel; diff --git a/psrdada_cpp/effelsberg/edd/detail/GatedSpectrometer.cu b/psrdada_cpp/effelsberg/edd/detail/GatedSpectrometer.cu index f32326097a9100ec512b4d7b43c75485814cf52c..7daa377a52ceac072855489d0732e5b81e1007f1 100644 --- a/psrdada_cpp/effelsberg/edd/detail/GatedSpectrometer.cu +++ b/psrdada_cpp/effelsberg/edd/detail/GatedSpectrometer.cu @@ -65,17 +65,14 @@ __global__ void update_baselines(float* __restrict__ baseLineG0, template <class HandlerType, class InputType, class OutputType> GatedSpectrometer<HandlerType, InputType, OutputType>::GatedSpectrometer( - const DadaBufferLayout &dadaBufferLayout, std::size_t selectedSideChannel, - std::size_t selectedBit, std::size_t fft_length, std::size_t naccumulate, - std::size_t nbits, HandlerType - &handler) : _dadaBufferLayout(dadaBufferLayout), - _selectedSideChannel(selectedSideChannel), _selectedBit(selectedBit), - _fft_length(fft_length), _naccumulate(naccumulate), + const GatedSpectrometerInputParameters &ip, HandlerType &handler) : _dadaBufferLayout(ip.dadaBufferLayout), + _selectedSideChannel(ip.selectedSideChannel), _selectedBit(ip.selectedBit), + _fft_length(ip.fft_length), _naccumulate(ip.naccumulate), _handler(handler), _fft_plan(0), _call_count(0), _nsamps_per_heap(4096) { // Sanity checks - assert(((nbits == 12) || (nbits == 8) || (nbits == 10))); + assert(((ip.nbits == 12) || (ip.nbits == 8) || (ip.nbits == 10))); assert(_naccumulate > 0); // check for any device errors @@ -92,16 +89,16 @@ GatedSpectrometer<HandlerType, InputType, OutputType>::GatedSpectrometer( << " output bit depth " << sizeof(IntegratedPowerType) * 8; assert((_dadaBufferLayout.getNSideChannels() == 0) || - (selectedSideChannel < _dadaBufferLayout.getNSideChannels())); // Sanity check of side channel value - assert(selectedBit < 64); // Sanity check of selected bit + (_selectedSideChannel < _dadaBufferLayout.getNSideChannels())); // Sanity check of side channel value + assert(_selectedBit < 64); // Sanity check of selected bit _nchans = _fft_length / 2 + 1; - inputDataStream = new InputType(fft_length, nbits, _dadaBufferLayout); + inputDataStream = new InputType(_fft_length, ip.nbits, _dadaBufferLayout); //How many output spectra per input block? - size_t nsamps_per_output_spectra = fft_length * naccumulate; + size_t nsamps_per_output_spectra = _fft_length * _naccumulate; size_t nsamps_per_pol = inputDataStream->getSamplesPerInputPolarization(); BOOST_LOG_TRIVIAL(debug) << "Samples per input polarization: " << nsamps_per_pol; @@ -126,7 +123,7 @@ GatedSpectrometer<HandlerType, InputType, OutputType>::GatedSpectrometer( // plan the FFT - size_t nsamps_per_buffer = _dadaBufferLayout.sizeOfData() * 8 / nbits; + size_t nsamps_per_buffer = _dadaBufferLayout.sizeOfData() * 8 / ip.nbits; int batch = nsamps_per_pol / _fft_length; int n[] = {static_cast<int>(_fft_length)}; BOOST_LOG_TRIVIAL(debug) << "Generating FFT plan:\n" diff --git a/psrdada_cpp/effelsberg/edd/src/GatedSpectrometer_cli.cu b/psrdada_cpp/effelsberg/edd/src/GatedSpectrometer_cli.cu index be8bed6c8ee845c5d73099dce261cae36ae87093..22d311931e167942cc8ae46069687f0ca47e17f2 100644 --- a/psrdada_cpp/effelsberg/edd/src/GatedSpectrometer_cli.cu +++ b/psrdada_cpp/effelsberg/edd/src/GatedSpectrometer_cli.cu @@ -24,60 +24,42 @@ const size_t ERROR_UNHANDLED_EXCEPTION = 2; } // namespace -struct GatedSpectrometerInputParameters -{ - effelsberg::edd::DadaBufferLayout dadaBufferLayout; - size_t selectedSideChannel; - size_t speadHeapSize; - size_t nSideChannels; - size_t selectedBit; - size_t fft_length; - size_t naccumulate; - unsigned int nbits; - std::string filename; - std::string output_type; -}; - template<typename T, class InputType, class OutputType > -void launchSpectrometer(const GatedSpectrometerInputParameters &i) +void launchSpectrometer(const effelsberg::edd::GatedSpectrometerInputParameters &i, const std::string &filename, const std::string &output_type) { MultiLog log("DadaBufferLayout"); - std::cout << "Running with output_type: " << i.output_type << std::endl; - if (i.output_type == "file") + std::cout << "Running with output_type: " << output_type << std::endl; + if (output_type == "file") { - SimpleFileWriter sink(i.filename); + SimpleFileWriter sink(filename); effelsberg::edd::GatedSpectrometer<decltype(sink), InputType, OutputType> - spectrometer(i.dadaBufferLayout, - i.selectedSideChannel, i.selectedBit, - i.fft_length, i.naccumulate, i.nbits, sink); + spectrometer(i, sink); DadaInputStream<decltype(spectrometer)> istream(i.dadaBufferLayout.getInputkey(), log, spectrometer); istream.start(); } - else if (i.output_type == "dada") + else if (output_type == "dada") { - DadaOutputStream sink(string_to_key(i.filename), log); - effelsberg::edd::GatedSpectrometer<decltype(sink), InputType, OutputType> spectrometer(i.dadaBufferLayout, - i.selectedSideChannel, i.selectedBit, - i.fft_length, i.naccumulate, i.nbits, sink); + DadaOutputStream sink(string_to_key(filename), log); + effelsberg::edd::GatedSpectrometer<decltype(sink), InputType, OutputType> + spectrometer(i, sink); DadaInputStream<decltype(spectrometer)> istream(i.dadaBufferLayout.getInputkey(), log, spectrometer); istream.start(); } - else if (i.output_type == "profile") + else if (output_type == "profile") { NullSink sink; - effelsberg::edd::GatedSpectrometer<decltype(sink), InputType, OutputType> spectrometer(i.dadaBufferLayout, - i.selectedSideChannel, i.selectedBit, - i.fft_length, i.naccumulate, i.nbits, sink); + effelsberg::edd::GatedSpectrometer<decltype(sink), InputType, OutputType> + spectrometer(i, sink); std::vector<char> buffer(i.dadaBufferLayout.getBufferSize()); cudaHostRegister(buffer.data(), buffer.size(), cudaHostRegisterPortable); @@ -97,12 +79,12 @@ void launchSpectrometer(const GatedSpectrometerInputParameters &i) } -template<typename T> void io_eval(const GatedSpectrometerInputParameters &inputParameters, const std::string &input_polarizations, const std::string &output_format) +template<typename T> void io_eval(const effelsberg::edd::GatedSpectrometerInputParameters &inputParameters, const std::string &input_polarizations, const std::string &output_format, const std::string &filename, const std::string &output_type) { if (input_polarizations == "Single" && output_format == "Power") { launchSpectrometer<T, effelsberg::edd::SinglePolarizationInput, - effelsberg::edd::GatedPowerSpectrumOutput>(inputParameters); + effelsberg::edd::GatedPowerSpectrumOutput>(inputParameters, filename, output_type); } else if (input_polarizations == "Dual" && output_format == "Power") { @@ -111,7 +93,7 @@ template<typename T> void io_eval(const GatedSpectrometerInputParameters &inputP else if (input_polarizations == "Dual" && output_format == "Stokes") { launchSpectrometer<T, effelsberg::edd::DualPolarizationInput, - effelsberg::edd::GatedFullStokesOutput>(inputParameters); + effelsberg::edd::GatedFullStokesOutput>(inputParameters, filename, output_type); } else { @@ -128,7 +110,7 @@ int main(int argc, char **argv) { try { key_t input_key; - GatedSpectrometerInputParameters ip; + effelsberg::edd::GatedSpectrometerInputParameters ip; std::time_t now = std::time(NULL); std::tm *ptm = std::localtime(&now); char default_filename[32]; @@ -136,7 +118,8 @@ int main(int argc, char **argv) { std::string input_polarizations = "Single"; std::string output_format = "Power"; - + std::string filename; + std::string output_type; /** Define and parse the program options */ namespace po = boost::program_options; @@ -150,11 +133,11 @@ int main(int argc, char **argv) { "The shared memory key for the dada buffer to connect to (hex " "string)"); desc.add_options()( - "output_type", po::value<std::string>(&ip.output_type)->default_value("file"), + "output_type", po::value<std::string>(&output_type)->default_value("file"), "output type [dada, file, profile]. Default is file. Profile executes the spectrometer 10x on random data and passes the ouput to a null sink." ); desc.add_options()( - "output_key,o", po::value<std::string>(&ip.filename)->default_value(default_filename), + "output_key,o", po::value<std::string>(&filename)->default_value(default_filename), "The key of the output bnuffer / name of the output file to write spectra " "to"); @@ -194,6 +177,10 @@ int main(int argc, char **argv) { desc.add_options()("naccumulate,a", po::value<size_t>(&ip.naccumulate)->required(), "The number of samples to integrate in each channel"); +// desc.add_options()("disable_gate,d", +// po::value<size_t>(&ip.disable_gate)->default_value(-1), +// "Disable processing of ND state 0,1. Select -1 (default) to process both."); +// desc.add_options()( "log_level", po::value<std::string>()->default_value("info")->notifier( [](std::string level) { set_log_level(level); }), @@ -215,9 +202,9 @@ int main(int argc, char **argv) { } po::notify(vm); - if (vm.count("output_type") && (!(ip.output_type == "dada" || ip.output_type == "file" || ip.output_type== "profile") )) + if (vm.count("output_type") && (!(output_type == "dada" || output_type == "file" || output_type== "profile") )) { - throw po::validation_error(po::validation_error::invalid_option_value, "output_type", ip.output_type); + throw po::validation_error(po::validation_error::invalid_option_value, "output_type", output_type); } if (vm.count("input_polarizations") && (!(input_polarizations == "Single" || input_polarizations == "Dual") )) @@ -248,7 +235,7 @@ int main(int argc, char **argv) { ip.dadaBufferLayout.intitialize(input_key, ip.speadHeapSize, ip.nSideChannels); - io_eval<float>(ip, input_polarizations, output_format); + io_eval<float>(ip, input_polarizations, output_format, filename, output_type ); } catch (std::exception &e) { std::cerr << "Unhandled Exception reached the top of main: " << e.what() diff --git a/psrdada_cpp/effelsberg/edd/test/src/GatedSpectrometerTest.cu b/psrdada_cpp/effelsberg/edd/test/src/GatedSpectrometerTest.cu index e352aea3580735c76d4d2d301417c97aa1610e48..b1cfbf77d771d17f1ecb74ba04937d8343a8a442 100644 --- a/psrdada_cpp/effelsberg/edd/test/src/GatedSpectrometerTest.cu +++ b/psrdada_cpp/effelsberg/edd/test/src/GatedSpectrometerTest.cu @@ -313,29 +313,29 @@ class GatedTestSinkSinglePol{ TEST(GatedSpectrometer, processingSinglePol) { - const size_t nbits = 8; const size_t nHeaps = 1024; - const size_t fft_length = 1024 * 64; - const size_t naccumulate = 4096 * nHeaps / fft_length; - const size_t heapSize = 4096 * nbits / 8; - const size_t inputBufferSize = nHeaps * (heapSize + 64 / 8); + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters ip; + ip.nbits = 8; + ip.fft_length = 1024 * 64; + ip.naccumulate = 4096 * nHeaps / ip.fft_length; + ip.selectedBit = 0; + ip.selectedSideChannel = 0; + ip.speadHeapSize = 4096 * ip.nbits / 8; + + const size_t inputBufferSize = nHeaps * (ip.speadHeapSize + 64 / 8); psrdada_cpp::DadaDB idbuffer(5, inputBufferSize, 1, 4096); idbuffer.create(); - psrdada_cpp::effelsberg::edd::DadaBufferLayout bufferLayout(idbuffer.key(), heapSize, 1); + ip.dadaBufferLayout = psrdada_cpp::effelsberg::edd::DadaBufferLayout(idbuffer.key(), ip.speadHeapSize, 1); - GatedTestSinkSinglePol sink(fft_length, nHeaps, naccumulate); + GatedTestSinkSinglePol sink(ip.fft_length, nHeaps, ip.naccumulate); psrdada_cpp::effelsberg::edd::GatedSpectrometer< GatedTestSinkSinglePol, psrdada_cpp::effelsberg::edd::SinglePolarizationInput, psrdada_cpp::effelsberg::edd::GatedPowerSpectrumOutput> - spectrometer(bufferLayout, - 0, 0, - fft_length, - naccumulate, nbits, - sink); + spectrometer(ip, sink); char *raw_buffer = new char[inputBufferSize]; @@ -402,29 +402,30 @@ class GatedTestSinkFullStokes{ TEST(GatedSpectrometer, processingFullStokes) { - const size_t nbits = 8; const size_t nHeaps = 1024; - const size_t fft_length = 1024 * 64; - const size_t naccumulate = 4096 * nHeaps / fft_length; - const size_t heapSize = 4096 * nbits / 8; - const size_t inputBufferSize = 2 * nHeaps * (heapSize + 64 / 8) ; + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters ip; + ip.nbits = 8; + ip.fft_length = 1024 * 64; + ip.naccumulate = 4096 * nHeaps / ip.fft_length; + ip.selectedBit = 0; + ip.selectedSideChannel = 0; + ip.speadHeapSize = 4096 * ip.nbits / 8; + + + const size_t inputBufferSize = 2 * nHeaps * (ip.speadHeapSize + 64 / 8) ; psrdada_cpp::DadaDB idbuffer(5, inputBufferSize, 1, 4096); idbuffer.create(); - psrdada_cpp::effelsberg::edd::DadaBufferLayout bufferLayout(idbuffer.key(), heapSize, 1); + psrdada_cpp::effelsberg::edd::DadaBufferLayout bufferLayout(idbuffer.key(), ip.speadHeapSize, 1); - GatedTestSinkFullStokes sink(fft_length, nHeaps, naccumulate); + GatedTestSinkFullStokes sink(ip.fft_length, nHeaps, ip.naccumulate); psrdada_cpp::effelsberg::edd::GatedSpectrometer< GatedTestSinkFullStokes, psrdada_cpp::effelsberg::edd::DualPolarizationInput, psrdada_cpp::effelsberg::edd::GatedFullStokesOutput> - spectrometer(bufferLayout, - 0, 0, - fft_length, - naccumulate, nbits, - sink); + spectrometer(ip, sink); char *raw_buffer = new char[inputBufferSize]; @@ -472,6 +473,19 @@ struct gated_params std::size_t nbits; std::size_t nHeaps; std::string msg; + + + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters getInputParams() + { + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters ip; + ip.nbits = nbits; + ip.fft_length = fft_length; + ip.naccumulate = naccumulate; + ip.selectedBit = 0; + ip.selectedSideChannel = 0; + ip.speadHeapSize = 4096 * nbits / 8; + return ip; + } }; @@ -479,22 +493,24 @@ class ExecutionTests: public testing::TestWithParam<gated_params> { // Test that the spectrometers execute without error in certain parameter // settings + + }; TEST_P(ExecutionTests, SinglePolOutput) { gated_params params = GetParam(); + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters ip = params.getInputParams(); - const size_t heapSize = 4096 * params.nbits / 8; - const size_t inputBufferSize = params.nHeaps * (heapSize + 64 / 8); + const size_t inputBufferSize = params.nHeaps * (ip.speadHeapSize + 64 / 8); psrdada_cpp::DadaDB buffer(5, inputBufferSize, 1, 4096); buffer.create(); - psrdada_cpp::effelsberg::edd::DadaBufferLayout bufferLayout(buffer.key(), heapSize, 1); + ip.dadaBufferLayout = psrdada_cpp::effelsberg::edd::DadaBufferLayout(buffer.key(), ip.speadHeapSize, 1); // Test buffer consistency with input parameters - EXPECT_EQ(bufferLayout.getNHeaps(), params.nHeaps); + EXPECT_EQ(ip.dadaBufferLayout.getNHeaps(), params.nHeaps); struct output_pointer { @@ -564,11 +580,7 @@ TEST_P(ExecutionTests, SinglePolOutput) TestSink, psrdada_cpp::effelsberg::edd::SinglePolarizationInput, psrdada_cpp::effelsberg::edd::GatedPowerSpectrumOutput> - spectrometer(bufferLayout, - 0, 0, - params.fft_length, - params.naccumulate, params.nbits, - sink); + spectrometer(ip, sink); char *raw_buffer = new char[inputBufferSize]; @@ -584,7 +596,7 @@ TEST_P(ExecutionTests, SinglePolOutput) // 3 Block, set value and SCI raw_buffer[122] = int8_t(13); - uint64_t* sc_items = reinterpret_cast<uint64_t*>(raw_buffer + params.nHeaps * heapSize); + uint64_t* sc_items = reinterpret_cast<uint64_t*>(raw_buffer + params.nHeaps * ip.speadHeapSize); for (size_t i = 0; i < params.nHeaps; i++) SET_BIT(sc_items[i], 0); EXPECT_NO_THROW(spectrometer(buff)) << params.msg; @@ -601,16 +613,16 @@ TEST_P(ExecutionTests, SinglePolOutput) TEST_P(ExecutionTests, FullStokesOutput) { gated_params params = GetParam(); + psrdada_cpp::effelsberg::edd::GatedSpectrometerInputParameters ip = params.getInputParams(); - const size_t heapSize = 4096 * params.nbits / 8; - const size_t inputBufferSize = params.nHeaps * (heapSize + 64 / 8); + const size_t inputBufferSize = params.nHeaps * (ip.speadHeapSize + 64 / 8); psrdada_cpp::DadaDB buffer(5, inputBufferSize, 1, 4096); buffer.create(); - psrdada_cpp::effelsberg::edd::DadaBufferLayout bufferLayout(buffer.key(), heapSize, 1); + ip.dadaBufferLayout = psrdada_cpp::effelsberg::edd::DadaBufferLayout(buffer.key(), ip.speadHeapSize, 1); // Test buffer consistency with input parameters - EXPECT_EQ(bufferLayout.getNHeaps(), params.nHeaps); + EXPECT_EQ(ip.dadaBufferLayout.getNHeaps(), params.nHeaps); psrdada_cpp::NullSink sink; @@ -618,11 +630,8 @@ TEST_P(ExecutionTests, FullStokesOutput) psrdada_cpp::NullSink, psrdada_cpp::effelsberg::edd::DualPolarizationInput, psrdada_cpp::effelsberg::edd::GatedFullStokesOutput> - spectrometer(bufferLayout, - 0, 0, - params.fft_length, - params.naccumulate, params.nbits, - sink); + spectrometer(ip, sink); + char *raw_buffer = new char[inputBufferSize]; memset(raw_buffer, 0, inputBufferSize);