diff --git a/src/ascot5_main.c b/src/ascot5_main.c index 219c6bfc..00cef459 100644 --- a/src/ascot5_main.c +++ b/src/ascot5_main.c @@ -64,6 +64,7 @@ #include "diag.h" #include "B_field.h" #include "plasma.h" +#include "rfof.h" #include "print.h" #include "simulate.h" #include "particle.h" @@ -233,9 +234,7 @@ int main(int argc, char** argv) { offload_free_offload(&offload_data, &offload_array, &int_offload_array); if(sim.enable_icrh) { - rfof_interface_deallocate_rfof_input_param( - &(sim.rfof_data.cptr_rfof_input_params)); - rfof_interface_deallocate_rfglobal(&(sim.rfof_data.cptr_rfglobal)); + rfof_free_offload(&sim.rfof_data); } /* Write output and clean */ diff --git a/src/libascot.c b/src/libascot.c index f6b456f0..191599d3 100644 --- a/src/libascot.c +++ b/src/libascot.c @@ -1015,7 +1015,7 @@ void libascot_eval_ratecoeff( * @param Eminus_imag Imaginary part of the right-handed electric field * component of the wave [V/m]. * @param res_cond value of the resonance condition where zero is the resonance - * [unitless]. + * [dimensionless]. */ void libascot_eval_rfof( sim_offload_data* sim_offload_data, real* B_offload_array, int Neval, @@ -1027,40 +1027,40 @@ void libascot_eval_rfof( B_field_init(&sim.B_data, &sim_offload_data->B_offload_data, B_offload_array); rfof_init(&sim.rfof_data, &sim_offload_data->rfof_data); - #pragma omp parallel for - for(int k = 0; k < Neval; k++) { - real B[3]; - if( B_field_eval_B(B, R[k], phi[k], z[k], t[k], &sim.B_data) ) { - continue; - } - real B_magn = sqrt(B[0]*B[0] + B[1]*B[1] + B[2]*B[2]); - real gyrofreq = q * B_magn / mass; - + #pragma omp parallel + { /* The function that evaluates resonance condition takes an RFOF marker - * as an input. However, only the R and vpar values are actually used. - * Therefore, we initialize a dummy marker and adjust only the values of - * R and vpar. */ - void* marker_pointer; + * as an input. However, only the R and vpar values are actually used. + * Therefore, we initialize a dummy marker and adjust only the values of + * R and vpar. */ + rfof_marker rfof_mrk; int dummy_int = 1; - real dummy_real = -999.0; - rfof_interface_allocate_rfof_marker(&marker_pointer); - rfof_interface_set_marker_pointers(&marker_pointer, &dummy_int, - &dummy_real, &(R[k]), &dummy_real, &dummy_real, &dummy_real, - &dummy_real, &dummy_real, &dummy_real, &dummy_real, &dummy_real, - &dummy_real, &vpar, &dummy_real, &gyrofreq, &dummy_real, - &dummy_real, &dummy_int, &dummy_int); - - int nharm; /* For storing return value which is not used */ - rfof_interface_eval_resonance_function( - &marker_pointer, &(sim.rfof_data.cptr_rfglobal), - &(res_cond[k]), &nharm); - - // TODO: this should return a non-zero value if the evaluation failed. - rfof_interface_get_rf_wave_local( - &(R[k]), &(z[k]), &dummy_real, &dummy_real, - &(sim.rfof_data.cptr_rfglobal), &(Eplus_real[k]), &(Eminus_real[k]), - &(Eplus_imag[k]), &(Eminus_imag[k])); - rfof_interface_deallocate_marker(&marker_pointer); - continue; + real dummy_real = 0.0; + rfof_set_up(&rfof_mrk, &sim.rfof_data); + + #pragma omp for + for(int k = 0; k < Neval; k++) { + real B[3]; + if( B_field_eval_B(B, R[k], phi[k], z[k], t[k], &sim.B_data) ) { + continue; + } + real B_magn = sqrt(B[0]*B[0] + B[1]*B[1] + B[2]*B[2]); + real gyrofreq = q * B_magn / mass; + rfof_set_marker_manually(&rfof_mrk, &dummy_int, + &dummy_real, &(R[k]), &dummy_real, &dummy_real, &dummy_real, + &dummy_real, &dummy_real, &dummy_real, &dummy_real, &dummy_real, + &dummy_real, &vpar, &dummy_real, &gyrofreq, &dummy_real, + &dummy_real, &dummy_int, &dummy_int); + + int nharm; /* For storing return value which is not used */ + rfof_eval_resonance_function( + &(res_cond[k]), &nharm, &rfof_mrk, &(sim.rfof_data.rfglobal)); + + // TODO: this should return a non-zero value for failed evaluations + rfof_eval_rf_wave( + &(Eplus_real[k]), &(Eminus_real[k]), &(Eplus_imag[k]), + &(Eminus_imag[k]), R[k], z[k], &sim.rfof_data); + } + rfof_tear_down(&rfof_mrk); } } diff --git a/src/rfof.c b/src/rfof.c index 1ba31b12..09897b4e 100644 --- a/src/rfof.c +++ b/src/rfof.c @@ -3,22 +3,32 @@ * @brief Contains the function to be called during the simulation when using * ICRH. Requires librfof.so library which contains the Fortran routines. **/ - #include #include #include #include #include "physlib.h" #include "particle.h" -#include "rfof_interface.h" - -#ifdef RFOF +#include "rfof.h" +/** + * @brief + */ +typedef struct rfof_output { + double dmu; /**< Change in magnetic moment due to ICRH kick. */ + double dvpar; /**< Change in parallel velocity component due to ICRH + kick. */ + double de; /**< Change in energy due to a single ICRH kick [J]. */ + double deCumulative; /**< Change in energy due to possibly several ICRH + kicks during an orbit time step [J] */ + double dpitch; /**< Change in pitch due to ICRH kick */ + double maxAcc; /**< Maximum acceleration allowed by RFOF */ + double RFdt; /**< time step suggested by RFOF */ +} rfof_output; -/* INITIALISATION */ - +#ifdef RFOF void __ascot5_icrh_routines_MOD_call_initev_excl_marker_stuff( - char* xml_filename, int **xml_filename_len,void** cptr_rfglobal, + char* xml_filename, int **xml_filename_len, void** cptr_rfglobal, void** cptr_rfof_input_params); void __ascot5_icrh_routines_MOD_call_initialise_res_mem(void** cptr_mem, int* cptr_mem_shape_i, int* cptr_mem_shape_j, void** cptr_rfglobal, @@ -26,13 +36,6 @@ void __ascot5_icrh_routines_MOD_call_initialise_res_mem(void** cptr_mem, void __ascot5_icrh_routines_MOD_call_initialise_diagnostics( void** cptr_RFglobal, void** cptr_diagno); -/* NOTE: There is no separate routine for initialising the markers; it is done - using the call_set_marker_pointers function with the argument - is_already_addlocated=0 */ - - -/* STUFF TO DO BETWEEN KICKS */ - void __ascot5_icrh_routines_MOD_call_set_marker_pointers(void** cptr_marker, int** id, real** weight, real** R, real** phi, real** z, real** psi, real** charge, real** mass, real** Ekin, real** velocity, real** mu, @@ -40,23 +43,14 @@ void __ascot5_icrh_routines_MOD_call_set_marker_pointers(void** cptr_marker, real** vdriftRho, real** acc, int* isOrbitTimeAccelerated, int* is_already_allocated); - -/* KICK */ - void __ascot5_icrh_routines_MOD_call_rf_kick(double*time, double*dtin, - int* myMPIprocID, prt_rfof* rfof_data_pack, void** cptr_marker, + int* mpi_rank, rfof_output* out, void** cptr_marker, void**cptr_mem, void** cptr_rfglobal, void** cptr_rfdiagno, int *err, int*mem_shape_i, int*mem_shape_j); - -/* RESET RESONANCE MEMORY */ - void __ascot5_icrh_routines_MOD_call_reset_res_mem(void** rfof_mem_pointer, int* mem_shape_i, int* mem_shape_j); - -/* DEALLOCATIONS*/ - void __ascot5_icrh_routines_MOD_call_deallocate_rfof_input_param( void** cptr_rfof_input_param); void __ascot5_icrh_routines_MOD_call_deallocate_rfglobal(void** cptr_rfglobal); @@ -65,58 +59,37 @@ void __ascot5_icrh_routines_MOD_call_deallocate_res_mem(void** cptr_res_mem, void __ascot5_icrh_routines_MOD_call_deallocate_diagnostics(void** cptr_diagno); void __ascot5_icrh_routines_MOD_deallocate_marker(void** cptr_rfof_marker); - -/* FOR VISUALISING ICRH WAVE FIELD */ - void __ascot5_icrh_routines_MOD_get_rf_wave_local_v2(real* R, real* z, - real* rho_tor, real* theta, void** cptr_wi, real* e_plus_real, + real* rho_tor, real* theta, void** cptr_rfglobal, real* e_plus_real, real* e_minus_real, real* e_plus_imag, real* e_minus_imag); void __ascot5_icrh_routines_MOD_eval_resonance_function(void** cptr_marker, void** cptr_rfglobal, real* omega_res, int* nharm); - -/* FOR DEBUGGING (n.b. TOTALLY OBSOLETE, OBVIOUSLY) */ - void __ascot5_icrh_routines_MOD_print_marker_stuff(void** marker_pointer); #endif - - - -/********************************* FUNCTIONS **********************************/ - - -/* INITIALISATION */ - /** - * @brief Initialise everyting excluding marker stuff. Reads the ICRH (RFOF) - * inputs (xml, xsd, ASCII) and initialises the wave field (variable name: - * RFglobal). + * @brief Initialise input data. * - * @param xml_filename Name of the xml file (less than 124 char) - * @param xml_filename_len Length of the xml_filename (excluding the '\0' at the - * end) - * @param cptr_rfglobal void pointer to the constructed wave field. Cannot be - * used to access the wave field but acts as a reference. - * @param cptr_rfof_input_params void pointer to an RFOF struct containing the - * input parameters. Only relevant when constructing the resonance memorys later - * on. -*/ -char xml_filename[128] = "rfof_codeparam.xml"; -void rfof_interface_initev_excl_marker_stuff(rfof_data* rfof_data) { - + * Reads the ICRH (RFOF) inputs (xml, xsd, ASCII) and initialises the wave + * field. + * + * Note: Despite being called init_offload, the offloading is not yet supported. + * The name implies where in the main program this should be called. + * + * @param rfof_data pointer to the RFOF data structure + */ +void rfof_init_offload(rfof_data* rfof_data) { #ifdef RFOF - //const char* xml_filename = RFOF_CODEPARAM_XML; + const char xml_filename[128] = RFOF_CODEPARAM_XML; int xml_filename_len = strlen(xml_filename); int*xml_filename_len_ptr = &xml_filename_len; - printf("xml_filename = %s\n", xml_filename); - printf("len(eml_file_name) = %d\n", xml_filename_len); __ascot5_icrh_routines_MOD_call_initev_excl_marker_stuff(xml_filename, - &xml_filename_len_ptr, &(rfof_data->cptr_rfglobal), - &(rfof_data->cptr_rfof_input_params)); + &xml_filename_len_ptr, &(rfof_data->rfglobal), + &(rfof_data->rfof_input_params)); #endif -}; +} /** * @brief Initialize RFOF data on target. @@ -128,289 +101,161 @@ void rfof_interface_initev_excl_marker_stuff(rfof_data* rfof_data) { * @param rfof_offload_data rfof data on host */ void rfof_init(rfof_data* rfof, rfof_data* rfof_offload_data) { - rfof->cptr_rfglobal = rfof_offload_data->cptr_rfglobal; - rfof->cptr_rfof_input_params = rfof_offload_data->cptr_rfof_input_params; + rfof->rfglobal = rfof_offload_data->rfglobal; + rfof->rfof_input_params = rfof_offload_data->rfof_input_params; } /** - * @brief Initialises resonance memory for rfof markers. To be called before the - * time step loop. (Each RFOF marker has its ows resonance memory matrix which - * in turn has elements corresponding to each wave and its mode.) - * @param cptr_mem Handle to corresponding Fortran resonance memory matrix - * pointer - * @param cptr_mem_shape_i Size of the first dimension of the resonance memory - * matrix - * @param cptr_mem_shape_j Size of the second dimension of the resonance memory - * matrix - * @param cptr_rfglobal Handle to the Fortan wave field struct - * @param cptr_rfof_input_param Handle to the Fortran input parameter struct - * */ -void rfof_interface_initialise_res_mem(void** cptr_mem, int* cptr_mem_shape_i, - int* cptr_mem_shape_j, void** cptr_rfglobal, void** cptr_rfof_input_param) { + * @brief Deallocates the rfof_input_param struct on the fortran side. There + * exists only one copy of this struct and therefore it is to be deallocated in + * the simulate.c after the loop is completed. + * @param cptr_rfof_input_param Handle to rfof input param struct on the Fortran + * side. + */ +void rfof_free_offload(rfof_data* rfof) { #ifdef RFOF - __ascot5_icrh_routines_MOD_call_initialise_res_mem(cptr_mem, - cptr_mem_shape_i, cptr_mem_shape_j, cptr_rfglobal, - cptr_rfof_input_param); + __ascot5_icrh_routines_MOD_call_deallocate_rfof_input_param( + &rfof->rfof_input_params); + __ascot5_icrh_routines_MOD_call_deallocate_rfglobal(&rfof->rfglobal); #endif -}; +} /** - * @brief Initialises rfof diagnostics. These are not used but are given as - * dummy inputs to the kick routine. To be called before the time step loop. - * @param cptr_rfglobal Handle to the Fortan wave field struct - * @param cptr_diagno Handle to the Fortran diagnostics struct -*/ -void rfof_interface_initialise_diagnostics(void** cptr_rfglobal, - void** cptr_diagno) { + * @brief Initialises resonance history, diagnostics, and the marker struct + * + * This function is to be called before the simulation loop. + * + * @param rfof_mrk pointer to the local RFOF marker data + * @param rfof pointer to the shared RFOF data + */ +void rfof_set_up(rfof_marker* rfof_mrk, rfof_data* rfof) { #ifdef RFOF - __ascot5_icrh_routines_MOD_call_initialise_diagnostics(cptr_rfglobal, - cptr_diagno); + for(int i=0; i< NSIMD; i++) { + /* Initialize marker data with dummy values */ + int is_already_allocated = 0; + __ascot5_icrh_routines_MOD_call_set_marker_pointers( + &(rfof_mrk->p[i]), NULL, NULL, NULL, NULL, NULL, NULL, NULL, NULL, + NULL, NULL, NULL, NULL, NULL, NULL, NULL, NULL, NULL, NULL, + &is_already_allocated); + + /* Initialize resonance history */ + __ascot5_icrh_routines_MOD_call_initialise_res_mem( + &rfof_mrk->history_array[i], &rfof_mrk->nrow[i], + &rfof_mrk->ncol[i], &rfof->rfglobal, + &rfof->rfof_input_params); + + /* Initialize diagnostics */ + __ascot5_icrh_routines_MOD_call_initialise_diagnostics( + &rfof->rfglobal, &(rfof_mrk->diag_array[i])); + } #endif -}; +} /** - * @brief Allocates memory for an rfof marker on the fortran side. To be called - * before the time step loop. The memory should be deallocated at the end by - * calling the corresponding deallocation routine at the end of this file. - * @param rfof_marker_pointer Handle to rfof marker struct. -*/ -void rfof_interface_allocate_rfof_marker(void** rfof_marker_pointer) { + * @brief Deallocates the data structs used by the RFOF marker simulation data + * + * This function is to be called after the simulation loop. + * + * @param rfof_mrk pointer to the RFOF marker simulation data + */ +void rfof_tear_down(rfof_marker* rfof_mrk) { #ifdef RFOF - - /* These are actually dummies and could be whatever you want */ - real dummy_real_rfof = 41.99; - real* dummy_real_rfof_ptr = &dummy_real_rfof; - int dummy_int_rfof = 42; - int *dummy_int_rfof_ptr = &dummy_int_rfof; - - /** @brief Needs to be zero. This is not stored in the marker; this input - * parameter is only used when calling the call_set_marker_pointers to - * determine whether to allocate new memory or use an existing marker */ - int is_already_allocated = 0; - - /* Allocates memory and sets some dummy values to the marker fields */ - __ascot5_icrh_routines_MOD_call_set_marker_pointers( - rfof_marker_pointer, - &dummy_int_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_real_rfof_ptr, - &dummy_int_rfof, /* isOrbitTimeAccelerated: Currently a dummy - (The fortran routine does nothing with this - and sets orbit_time_accelerated to false - regardless) */ - &is_already_allocated); + for(int i=0; i< NSIMD; i++) { + __ascot5_icrh_routines_MOD_call_deallocate_res_mem( + &rfof_mrk->history_array[i], &rfof_mrk->nrow[i], + &rfof_mrk->ncol[i]); + __ascot5_icrh_routines_MOD_call_deallocate_diagnostics( + &rfof_mrk->diag_array[i]); + __ascot5_icrh_routines_MOD_deallocate_marker(&rfof_mrk->p[i]); + } #endif -}; - - -/* STUFF TO DO BETWEEN KICKS */ +} /** - * @brief Sets the RFOF marker's fields (all Fortran pointers). If the RFOF - * marker has not been allocated previously, this routine can be used to reate a - * marker for the RFOF routines on the fortran side. This allocation of a new - * RFOF marker should be done using the function - * rfof_interface_allocate_rfof_marker defined in this file. - * @param cptr_marker Handle to the RFOF marker struct in Fortran - * @param id Id of the marker (unclear whether truly needed) - * @param weight Marker weight - * @param R Major radius co-ordinate - * @param phi Toroidal angle co-ordinate - * @param z Vertical co-ordinate - * @param psi Poloidal flux function (following the ITM and ITER conventions; - * COCOS 13/11; not divided by \f$ 2\pi \f$) - * @param charge Charge of the particles represented by the marker. - * @param mass Mass of the particles represented by the marker - * @param Ekin Kinetic energy - * @param Velocity Speed of the particle. Obviously. - * @param mu Magnetic moment - * @param pphicanonical Canonical momentum conjugate to phi - * @param vpar Component of velocity parallel to B - * @param vperp Magnitude of velocity perpendicular to B - * @param gyrof Gyrofrequency - * @param vdriftRho The component of the drift velocity w.r.t. the radial - * direction (unclear if truly needed). This can used at least when evaluating - * the Doppler shift in the resonance condition. - * @param acc Time acceleration factor (not in use anymore, should be 1; NOT 0) - * @param isOrbitTimeAcclerated False as implied above in acc. Logical in - * fortran; represented by int (4 bytes) in C. Currently the Fortran routine - * does nothing with this parameter but instead sets it to false for all - * markers. - * @param is_already_allocated If true, the routine tries to fetch the existing - * particle corresponding to cptr_marker. If false, the routine allocates - * memory for an RFOF marker and returns the c_location of that fortran struct - * in cptr_marker. -*/ -void rfof_interface_set_marker_pointers(void** cptr_marker, int* id, - real* weight, real* R, real* phi, real* z, real* psi, real* charge, - real* mass, real* Ekin, real* velocity, real* mu, real* pphicanonical, - real* vpar, real* vperp, real* gyrof, real* vdriftRho, real* acc, - int* isOrbitTimeAccelerated, int* is_already_allocated) { + * @brief Clears resonance history of an RFOF marker + * + * History should be cleared whenever a marker finishes simulation. The new + * marker cannot receive ICRH kicks during the first two time steps as its + * resonance history must have at least two data points stored to estimate + * the resonance location. + * + * @param rfof_mrk pointer to the RFOF marker simulation data + * @param i index in the NSIMD array for the marker whose history is cleared + */ +void rfof_clear_history(rfof_marker* rfof_mrk, int i) { #ifdef RFOF - __ascot5_icrh_routines_MOD_call_set_marker_pointers(cptr_marker, &id, - &weight, &R, &phi, &z, &psi, &charge, &mass, &Ekin, &velocity, &mu, - &pphicanonical, &vpar, &vperp, &gyrof, &vdriftRho, &acc, - isOrbitTimeAccelerated, is_already_allocated); + __ascot5_icrh_routines_MOD_call_reset_res_mem( + &rfof_mrk->history_array[i], &rfof_mrk->nrow[i], &rfof_mrk->ncol[i]); #endif -}; - +} /** - * @brief Function to be called in the main simulation loop during each step - * when following the guiding centre. + * @brief Check if the marker is in resonance and apply kick * * 1. Updates the fields of the rfof_marker based on the given input * ascot_marker. + * * 2. Calls the "kick" function, which * a) Checks resonance condition and * b) if in resonance, kicks marker and updates velocity of the rfof marker * and consequently also ascot marker (as only pointers are passed when * creating the rfof marker). - * If the proposed time step, hin, was too large, hout_rfof will be negative - * indicating that the whole time step has to be redone with a smaller time - * step. - * @param ascot_marker Ascot marker (no points for those who guessed it) - * @param hin Time step proposed by the simulation loop - * @param hout_rfof If negative, ICRH kick failed because hin was too large. In - * that case the absolute value of hout_rfof should be used when redoing the - * simulation loop. For successfull steps, this is the estimate to the next - * resonance. - * @param rfof_data A "package" of two ICRH related void pointers to Fortran - * routines. Out of the two, only the pointer to the global wave field is used. - * This variable is not to be confused with the local variable rfof_data_pack - * defined inside this function. - * @param Bdata The magnetic field. Needed for evaluating psi. - * @param rfof_marker_pointer_array Contains void pointers which are handles to - * rfof markers on the Fortran side. - * @param rfof_mem_pointer_array Contains void pointers which are handles to - * rfof markers' resonance memory matrices. - * @param rfof_diag_pointer_array Contains void pointers which are handles to - * rfof diagnostics structs (not in actual use but must be included). - * @param mem_shape_i Array of rfof resonance memory matrix's first dimensions. - * @param mem_shape_j Array of rfof resonance memory matrix's second dimensions. -*/ -void rfof_interface_do_rfof_stuff_gc(particle_simd_gc* ascot_marker, real* hin, - real* hout_rfof, rfof_data rfof_data, B_field_data* Bdata, - void** rfof_marker_pointer_array, void** rfof_mem_pointer_array, - void** rfof_diag_pointer_array, int* mem_shape_i, int* mem_shape_j) { + * + * The time step can fail if the marker overshoots the resonance. + * + * @param p pointer to marker simulation struct + * @param hin current time step + * @param hout suggestion for the next time step with negative sign + * indicating a failed step + * @param rfof_mrk pointer to the rfof marker simulation struct + * @param rfof_data pointer to the shared rfof data + * @param Bdata pointer to the magnetic field data needed to evaluate psi + */ +void rfof_resonance_check_and_kick_gc( + particle_simd_gc* p, real* hin, real* hout, rfof_marker* rfof_mrk, + rfof_data* rfof_data, B_field_data* Bdata) { #ifdef RFOF - for (int i=0; iid[i] > 0 && ascot_marker->running[i]) { - /*Some of the fields needed by the RFOF_marker struct are not - present in the ascot_marker struct and thus need to be evaluated - first. */ - - /** @brief Poloidal flux function (following the ITM and ITER - * conventions; COCOS 13/11; not divided by \f$ 2\pi \f$) */ - real psi; - - B_field_eval_psi(&psi, ascot_marker->r[i], ascot_marker->phi[i], - ascot_marker->z[i], ascot_marker->time[i], Bdata); - - /** @brief Norm of B field */ - real B = sqrt(pow(ascot_marker->B_r[i],2) + - pow(ascot_marker->B_phi[i],2) + pow(ascot_marker->B_z[i],2)); - - /** @brief Lorentz factor.*/ - real gamma = physlib_gamma_ppar(ascot_marker->mass[i], - ascot_marker->mu[i], ascot_marker->ppar[i], B); - - /** @brief Energy [J]*/ - real Ekin = physlib_Ekin_gamma(ascot_marker->mass[i], gamma); - - /** @brief Speed */ - real speed = physlib_vnorm_gamma(gamma); - - /** @brief Canonical momentum conjugate to phi. - * Should be in SI units */ - real p_phi = phys_ptoroid_gc(ascot_marker->charge[i], - ascot_marker->r[i], ascot_marker->ppar[i], psi, B, - ascot_marker->B_phi[i]); - - /** @brief pitch */ - real xi = physlib_gc_xi(ascot_marker->mass[i], ascot_marker->mu[i], - ascot_marker->ppar[i], B); - - /** @brief Parallel momentum */ - real v_par = speed*xi; - - /** @brief Perpendicular speed */ - real v_perp = phys_vperp_gc(speed, v_par); - - /** @brief Gyrofrequency */ - real gyrof = phys_gyrofreq_ppar(ascot_marker->mass[i], - ascot_marker->charge[i], ascot_marker->mu[i], - ascot_marker->ppar[i], B); - - /** @brief Velocity drift in the direction of rho. Used when - * evaluating the Doppler shift term in the resonance condition. */ - real vdriftRho = 0; - - /** @brief Time acceleration factor. EVEN IF TIME ACCELERATION IS - * OFF, THIS IS NOT A DUMMY BUT MUST BE SET TO 1.0 */ - real acc = 1.0; - - /** @brief Time accelecation on/off. NOTE: Fortran logical - * corresponds to C int (4 bytes). NOTE2: Obsolete at the moment, as - * the Fortran routine currently forces this to be false (0) - * regardless. */ - int isOrbitTimeAccelerated = 0; - - /** @brief Must be 1 (true) at this point as the marker is already - * allocated.*/ - int is_already_allocated = 1; - - - /* Now when we have all the needed parameters we call the rfof - routine which sets the pointers of the rfof_marker (on the Fortran - side. The situation is somewhat comparable to ordering pizza/kebab - from Kontula: it's convenient that the place exists but you'd rather - not go there yourself. We're in luck, though, as we happen to have a - courier.) */ - - int dummy_Id = 3141592; - int* dummy_Id_ptr = &dummy_Id; - - /* These fields needed by RFOF marker we can pass straight away */ - - /** @brief The weight is only used for RFOF's own diagnostics which - * are not of interest to us. Thusly, the weight given to the RFOF - * marker could be arbitrary; from a certain point of view it's a - * dummy. In other words, these are not the weights you are looking - * for.*/ - real* weight_ptr = &(ascot_marker->weight[i]); - real* r_ptr = &(ascot_marker->r[i]); - real* phi_ptr = &(ascot_marker->phi[i]); - real* z_ptr = &(ascot_marker->z[i]); - real* charge_ptr = &(ascot_marker->charge[i]); - real* mass_ptr = &(ascot_marker->mass[i]); - real* mu_ptr = &(ascot_marker->mu[i]); - - /* These fields needed to be evaluated inside this function. They - are not independent variables. */ + for(int i=0; iid[i] > 0 && p->running[i]) { + /* Evaluate derived quantities needed by librfof */ + real psi, B, gamma, Ekin, vnorm, P_phi, xi, v_par, v_perp, gyrof; + B_field_eval_psi(&psi, p->r[i], p->phi[i], p->z[i], p->time[i], + Bdata); + B = math_normc(p->B_r[i], p->B_phi[i], p->B_z[i]); + gamma = physlib_gamma_ppar(p->mass[i], p->mu[i], p->ppar[i], B); + Ekin = physlib_Ekin_gamma(p->mass[i], gamma); + vnorm = physlib_vnorm_gamma(gamma); + P_phi = phys_ptoroid_gc(p->charge[i], p->r[i], p->ppar[i], psi, B, + p->B_phi[i]); + xi = physlib_gc_xi(p->mass[i], p->mu[i], p->ppar[i], B); + v_par = vnorm*xi; + v_perp = phys_vperp_gc(vnorm, v_par); + gyrof = phys_gyrofreq_ppar(p->mass[i], p->charge[i], p->mu[i], + p->ppar[i], B); + real vdriftRho = 0; // Assuming this is not needed in librfof + real acceleration = 1.0; + int is_accelerated = 0; + int is_preallocated = 1; + + int dummy_id = -1; + int* dummy_Id_ptr = &dummy_id; + + real* weight_ptr = &(p->weight[i]); + real* r_ptr = &(p->r[i]); + real* phi_ptr = &(p->phi[i]); + real* z_ptr = &(p->z[i]); + real* charge_ptr = &(p->charge[i]); + real* mass_ptr = &(p->mass[i]); + real* mu_ptr = &(p->mu[i]); real* Ekin_ptr = &Ekin; real* psi_ptr = ψ - real* speed_ptr = &speed; - real* p_phi_ptr = &p_phi; + real* speed_ptr = &vnorm; + real* P_phi_ptr = &P_phi; real* v_par_ptr = &v_par; real* v_perp_ptr = &v_perp; real* gyrof_ptr = &gyrof; real* vdriftRho_ptr = &vdriftRho; - real* acc_ptr = &acc; + real* acc_ptr = &acceleration; /* Store the old value to be able to evaluate the new ppar after kick. */ @@ -418,12 +263,9 @@ void rfof_interface_do_rfof_stuff_gc(particle_simd_gc* ascot_marker, real* hin, /* Update the fields of RFOF marker */ __ascot5_icrh_routines_MOD_call_set_marker_pointers( - &rfof_marker_pointer_array[i], /* Note that the pointer to the - pointer is passed. */ + &(rfof_mrk->p[i]), &(dummy_Id_ptr), - &(weight_ptr), /* Number of real particles - represented by the marker. - Acts as a dummy. */ + &(weight_ptr), &(r_ptr), &(phi_ptr), &(z_ptr), @@ -433,22 +275,18 @@ void rfof_interface_do_rfof_stuff_gc(particle_simd_gc* ascot_marker, real* hin, &(Ekin_ptr), &(speed_ptr), &(mu_ptr), - &(p_phi_ptr), /* pphicanonical */ + &(P_phi_ptr), &(v_par_ptr), &(v_perp_ptr), &(gyrof_ptr), - &(vdriftRho_ptr), /* vdriftRho; EVALUATE IF ACTUALLY - NEEDED */ + &(vdriftRho_ptr), &(acc_ptr), + &is_accelerated, + &is_preallocated); - &isOrbitTimeAccelerated, /**< Int in C, logical in Fortran */ - &is_already_allocated); /**< Int in C, logical in Fortran */ - - - /** @brief Used for storing the "results" of calling RF kick, only - * the RFdt field is utilised as it returns the time step - * recommended by RFOF */ - prt_rfof rfof_data_pack = { + /* Contains the change in physical quantities and the suggested + * time-step after the kick */ + rfof_output rfof_data_pack = { .dmu = 0.0, .dvpar = 0.0, .de = 0.0, @@ -458,232 +296,116 @@ void rfof_interface_do_rfof_stuff_gc(particle_simd_gc* ascot_marker, real* hin, .RFdt = 0.0, }; - int err = 0; /**< "empty" input to kick */ - - /** @brief Number used to identify MPI nods during parallel - * execution. */ - int mpiprocid = 0; - - /* Ready to kick some ash */ - __ascot5_icrh_routines_MOD_call_rf_kick(&(ascot_marker->time[i]), - &(hin[i]), &mpiprocid, &rfof_data_pack, - &(rfof_marker_pointer_array[i]), &(rfof_mem_pointer_array[i]), - &(rfof_data.cptr_rfglobal), &(rfof_diag_pointer_array[i]), &err, - &(mem_shape_i[i]), &(mem_shape_j[i])); - - /* Some of the rfof marker's pointers where pointing to the fields - of the ascot marker but some are pointing to the local variables - inside this function (e.g. Ekin). We now need to update the rest of - the ASCOT marker fields (= ppar) accordingly.*/ - - - /* Update the parallel momentum of ASCOT marker after RF kick */ - ascot_marker->ppar[i] = ascot_marker->mass[i]*(v_par_old + - rfof_data_pack.dvpar); - - // The other proposed methods for evaluating ppar: - /* - printf("\nCompute ppar using different input fields of RFOF marker" - "to check consistency\n"); + int err = 0; + int mpi_rank = 0; // RFOF does not work with MPI yet - // 0. ppar_old just for comparison - printf("For reference, ppar_old = %e\n", ppar_old); + /* Ready to kick some ash (if in resonance) */ + __ascot5_icrh_routines_MOD_call_rf_kick( + &(p->time[i]), &(hin[i]), &mpi_rank, &rfof_data_pack, + &(rfof_mrk->p[i]), &(rfof_mrk->history_array[i]), + &(rfof_data->rfglobal), &(rfof_mrk->diag_array[i]), &err, + &(rfof_mrk->nrow[i]), &(rfof_mrk->ncol[i])); - // 1. Updata pitch and use pitch and v to get vpar. Then use ppar = - //m*vpar - xi = xi + rfof_data_pack.dpitch; - real vpar1 = xi*speed; - printf("1. ppar from v and pitch + dpitch = %e\n", - ascot_marker->mass[i]*vpar1); + /* Most marker phase-space coordinates are updated automatically + * via the pointers in rfof_mrk except ppar which we update here */ + p->ppar[i] = p->mass[i]*(v_par_old + rfof_data_pack.dvpar); - // 2. Get ppar from v_par and m - printf("2. ppar = m*vpar = %e\n", ascot_marker->mass[i]*v_par); - - // 3. Get vpar from speed and vperp - real vpar2 = sqrt(speed*speed - v_perp*v_perp); - printf("3. abs(ppar) from speed and vperp = %e\n", - vpar2*ascot_marker->mass[i]); - - // 4. From Ekin get speed, from mu get vperp, then get vpar and ppar - real speed2 = sqrt(2.0*Ekin/ascot_marker->mass[i]); - real vperp2 = sqrt(2.0*ascot_marker->mu[i]*B/ascot_marker->mass[i]); - real vpar3 = sqrt(speed2*speed2 - vperp2*vperp2); - printf("4. abs(ppar) from Ekin and mu = %e\n", - ascot_marker->mass[i]*vpar3); - - // 5. Get ppar from p_phi - real ppar5 = B/(ascot_marker->r[i]*ascot_marker->B_phi[i])* - (p_phi - ascot_marker->charge[i]*psi); - printf("5. ppar from p_phi = %e\n", ppar5); - - // 6. Get ppar from gyrof and mu - real ppar6 = ascot_marker->mass[i]*CONST_C*sqrt(( - ascot_marker->charge[i]*B/(ascot_marker->mass[i]*gyrof))* - (ascot_marker->charge[i]*B/(ascot_marker->mass[i]*gyrof)) - - 2*ascot_marker->mu[i]*B/(ascot_marker->mass[i]*CONST_C2) - 1.0); - printf("6. abs(ppar) from gyrof and mu = %e\n", ppar6); - - // 7. Get ppar from vpar_old and dvpar in rfof_data_pack - printf("7. ppar from vpar_old and dvpar in rfof_data_pack = %e\n", - ascot_marker->mass[i]*(v_par_old + rfof_data_pack.dvpar)); - */ - - - /* Check if dt was sufficiently small and assign hout_rfof - accordingly. */ if (err == 7) { - /*Interaction failed, particle overshot the resonance. Make - hout_rfof negative. The absolute value of rfof_hout in this case - is the estimate of what hin should have been so that the - resonance would not have been missed (obviously less than hin). - */ - hout_rfof[i] = -rfof_data_pack.RFdt; + /* Overshot the resonance. Mark the suggested time-step as + * negative to inform ASCOT to retry the time step */ + hout[i] = -rfof_data_pack.RFdt; } else { - /* Interaction was successful. dt returned by rfof is now the - estimate to the next resonance. */ - if(rfof_data_pack.RFdt == 0) { - /* This if is only for debugginf purposes when it is - possible that the kick is not called. */ - printf("rfof_datapack.RFdt = 0, " - "setting rf return dt to 1\n"); - hout_rfof[i] = 1; - } else { - /* This is where you normally go when you call kick and - there is no error. */ - hout_rfof[i] = rfof_data_pack.RFdt; - } + /* Interaction was successful. The suggested time-step is + * a guess how long till the marker enters the resonance */ + hout[i] = rfof_data_pack.RFdt; } - - }; - }; -#endif -}; - - -/* RESET RESONANCE MEMORY */ - -/** - * @brief Resets resonance memory of ICRH (RFOF) markers. Should be done when - * the marker dies and a new one is born. Note that a marker with a newly - * allocated or reseted memory cannot receive ICRH kicks during the first two - * time steps as its resonance memory must have at least two data points stored - * for it to be kicked. - * @param rfof_mem_pointer Handle to rfof resonance memory matrix. - * @param mem_shape_i RFOF resonance memory matrix's first dimension. - * @param mem_shape_j RFOF resonance memory matrix's second dimension. - */ -void rfof_interface_reset_icrh_mem(void** rfof_mem_pointer, int* mem_shape_i, - int* mem_shape_j) { -#ifdef RFOF - __ascot5_icrh_routines_MOD_call_reset_res_mem(rfof_mem_pointer, mem_shape_i, - mem_shape_j); -#endif -}; - - -/* DEALLOCATION ROUTINES */ - -/** - * @brief Deallocates the rfof_input_param struct on the fortran side. There - * exists only one copy of this struct and therefore it is to be deallocated in - * the simulate.c after the loop is completed. - * @param cptr_rfof_input_param Handle to rfof input param struct on the Fortran - * side. - */ -void rfof_interface_deallocate_rfof_input_param(void** cptr_rfof_input_param) { -#ifdef RFOF - __ascot5_icrh_routines_MOD_call_deallocate_rfof_input_param( - cptr_rfof_input_param); -#endif -}; - -/** - * @brief Deallocates the rfglobal struct (wave field) on the fortran side. - * There exists only one copy of this struct and therefore it is to be - * deallocated in the simulate.c after the loop is completed -- much like the - * input_param struct. - * @param cptr_rfglobal Handle to rfof wave field on the Fortran side. - */ -void rfof_interface_deallocate_rfglobal(void** cptr_rfglobal) { -#ifdef RFOF - __ascot5_icrh_routines_MOD_call_deallocate_rfglobal(cptr_rfglobal); + } + } #endif -}; +} /** - * @brief Deallocates the resonance memory matrix of a particle. To be done when - * the simulation has finished. If a marker dies but there are still new ones in - * the queue, the resonance memory matrix should only be resetted - * (see:rfof_interface_reset_icrh_mem), not deallocated. - * @param rfof_mem_pointer Handle to rfof resonance memory. - * @param mem_shape_i RFOF resonance memory matrix's first dimension. - * @param mem_shape_j RFOF resonance memory matrix's second dimension. + * @brief Explicitly set the coordinates in a marker struct + * + * This routine is only used for testing and in libascot. Only the first marker + * in the NSIMD array of rfof_marker is manipulated. + * + * @param rfof_mrk pointer to the RFOF marker data + * @param id marker identifier + * @param weight marker weight [prt/s] + * @param R major radius [m] + * @param phi toroidal angle [rad] + * @param z z coordinate [m] + * @param psi poloidal flux [Wb/rad] + * @param charge particle charge [C] + * @param mass particle mass [kg] + * @param Ekin kinetic energy [J] + * @param vnorm velocity [m/s] + * @param mu magnetic moment [J/T] + * @param Pphi canonical toroidal angular momentum [kg*m/s] + * @param vpar perpendicular velocity [m/s] + * @param vperp parallel velocity [m/s] + * @param gyrof gyrofrequency [rad/s] + * @param vdriftRho drift velocity in radial direction + * @param acc accleration factor, should be 1.0 always + * @param is_accelerated flag indicating whether acceleration is used, false + * always + * @param is_already_allocated flag whether to allocate a new marker struct, + * should be false always */ -void rfof_interface_deallocate_res_mem(void** cptr_res_mem, - int* cptr_mem_shape_i, int* cptr_mem_shape_j) { -#ifdef RFOF - __ascot5_icrh_routines_MOD_call_deallocate_res_mem(cptr_res_mem, - cptr_mem_shape_i, cptr_mem_shape_j); -#endif -}; - -/** - * @brief Deallocates the (dummy) dianostics of rfof markers. - * @param cptr_diagno Handle to rfof diagnostics struct. - */ -void rfof_interface_deallocate_diagnostics(void** cptr_diagno) { -#ifdef RFOF - __ascot5_icrh_routines_MOD_call_deallocate_diagnostics(cptr_diagno); -#endif -}; - -/** - * @brief Deallocates the rfof marker. - * @param cptr_rfof_marker Handle to rfof marker. - */ -void rfof_interface_deallocate_marker(void** cptr_rfof_marker) { +void rfof_set_marker_manually(rfof_marker* rfof_mrk, int* id, + real* weight, real* R, real* phi, real* z, real* psi, real* charge, + real* mass, real* Ekin, real* vnorm, real* mu, real* Pphi, + real* vpar, real* vperp, real* gyrof, real* vdriftRho, real* acc, + int* is_accelerated, int* is_already_allocated) { #ifdef RFOF - __ascot5_icrh_routines_MOD_deallocate_marker(cptr_rfof_marker); + __ascot5_icrh_routines_MOD_call_set_marker_pointers( + &rfof_mrk->p[0], &id, &weight, &R, &phi, &z, &psi, &charge, &mass, + &Ekin, &vnorm, &mu, &Pphi, &vpar, &vperp, &gyrof, &vdriftRho, &acc, + is_accelerated, is_already_allocated); #endif -}; - - -/* FOR VISUALISING ICRH WAVE FIELD */ +} /** - * @brief Return the local E+ and E- values of the ICRH field, given the - * coordinates. - * @param R Major radius - * @param z Vertical co-ordinate - * @param rho_tor - * @param theta - * @param cptr_rfglobal Void pointer to the RFglobal global wave field in - * fortran + * @brief Calculate the local E+ and E- values of the ICRH field + * * @param e_plus_real Re(E+) component of the local wave field * @param e_minus_real Re(E-) component of the local wave field * @param e_plus_imag Im(E+) component of the local wave field * @param e_minus_imag Im(E-) component of the local wave field + * @param R major radius coordinate [m] + * @param z z-coordinate [m] + * @param rfof pointer to the RFOF data structure */ -void rfof_interface_get_rf_wave_local(real* R, real* z, real* rho_tor, - real* theta, void** cptr_wi, real* e_plus_real, real* e_minus_real, - real* e_plus_imag, real* e_minus_imag) { +void rfof_eval_rf_wave( + real* e_plus_real, real* e_minus_real, real* e_plus_imag, + real* e_minus_imag, real R, real z, rfof_data* rfof) { #ifdef RFOF - __ascot5_icrh_routines_MOD_get_rf_wave_local_v2(R, z, rho_tor, theta, - cptr_wi, e_plus_real, e_minus_real, e_plus_imag, e_minus_imag); + real rho_tor, theta; // Apparently not used in librfof + __ascot5_icrh_routines_MOD_get_rf_wave_local_v2( + &R, &z, &rho_tor, &theta, &rfof->rfglobal, e_plus_real, e_minus_real, + e_plus_imag, e_minus_imag); #endif -}; +} /** - * @brief Function for evaluating the value of resonance function (0=resonance) + * @brief Evaluate the value of resonance function (zero at the resonance) + * + * This function finds the closest resonance and in addition to evaluating the + * resonance function it also returns the corresponding harmonic value. + * + * The resonance is evaluated for the first marker in the NSIMD array in + * rfof_marker fields. + * * @param cptr_marker void pointer to the rfof_marker - * @param cptr_rfglobal void pointer to the rfof wave field - * @param omega_res value of the resonance function - * @param nharm harmonic index + * @param rfof_data pointer to the RFOF data structure + * @param omega_res evaluated value of the resonance function + * @param nharm the number of the closest harmonic found */ -void rfof_interface_eval_resonance_function(void** cptr_marker, - void** cptr_rfglobal, real* omega_res, int* nharm){ +void rfof_eval_resonance_function( + real* omega_res, int* nharm, rfof_marker* rfof_mrk, rfof_data* rfof) { #ifdef RFOF - __ascot5_icrh_routines_MOD_eval_resonance_function(cptr_marker, - cptr_rfglobal, omega_res, nharm); + __ascot5_icrh_routines_MOD_eval_resonance_function( + &rfof_mrk->p[0], &rfof->rfglobal, omega_res, nharm); #endif -}; +} diff --git a/src/rfof.h b/src/rfof.h index da8ec022..0a556617 100644 --- a/src/rfof.h +++ b/src/rfof.h @@ -1,101 +1,68 @@ /** - * @file rfof_interface.h + * @file rfof.h * @brief Contains the functions to be called from the simulation loop when * using ICRH. **/ - -#ifndef RFOF_INTERFACE_H -#define RFOF_INTERFACE_H - #include +#include "ascot5.h" + +#ifndef RFOF_H +#define RFOF_H /** - * @brief A one-to-one copy of the same struct from ASCOT4. Passed to RFOF kick - * routine which stores some of the results in these fields. + * @brief Reusable struct for storing marker specific data during the simulation + * loop. + * + * The data in this struct is altered during the simulation. Only pointers are + * stored as the actual data is stored on the Fortran side. + */ +typedef struct rfof_marker { + void* p[NSIMD]; /**< The marker struct in a format required by librfof */ + void* history_array[NSIMD]; /**< Stores values of the resonance function for + estimating the next time-step */ + void* diag_array[NSIMD]; /**< C equivalents of Fortran diagnostics pointers + which are required but unused at the moment */ + int nrow[NSIMD]; /**< Number of rows in an resonance history matrix */ + int ncol[NSIMD]; /**< Number of columns in an resonance history matrix */ + +} rfof_marker; + +/** @brief RFOF simulation input data + * + * Immutable input data shared between all markers. The actual data is stored + * in the Fortran side and this struct only stores the pointers. */ -typedef struct prt_rfof { - double dmu; /**< Change in magnetic moment due to ICRH kick. */ - double dvpar; /**< Change in parallel velocity component due to ICRH - kick. */ - double de; /**< Change in energy due to a single ICRH kick [J]. */ - double deCumulative; /**< Change in energy due to possibly several ICRH - kicks during an orbit time step [J] */ - double dpitch; /**< Change in pitch due to ICRH kick */ - double maxAcc; /**< Maximum acceleration allowed by RFOF */ - double RFdt; /**< time step suggested by RFOF */ -} prt_rfof; - -/** @brief Struct containing void pointers to RFOF structs */ typedef struct { - void* cptr_rfof_input_params; /**< Pointer to rfof_input_param struct on - the fortran side */ - void* cptr_rfglobal; /**< Wave field; same for all markers */ + void* rfof_input_params; /**< Pointer to rfof_input_param struct on + the fortran side */ + void* rfglobal; /**< Wave field; same for all markers */ } rfof_data; - - - -/********************************* FUNCTIONS **********************************/ - - -/* INITIALISATION */ - -void rfof_interface_initev_excl_marker_stuff(rfof_data* rfof_data); +void rfof_init_offload(rfof_data* rfof_data); void rfof_init(rfof_data* rfof, rfof_data* rfof_offload_data); -void rfof_interface_initialise_res_mem(void** cptr_mem, int* cptr_mem_shape_i, - int* cptr_mem_shape_j, void** cptr_rfglobal, void** cptr_rfof_input_param); - -void rfof_interface_initialise_diagnostics(void** cptr_RFglobal, - void** cptr_diagno); - -void rfof_interface_allocate_rfof_marker(void** rfof_marker_pointer); - +void rfof_free_offload(rfof_data* rfof); -/* STUFF TO DO BETWEEN KICKS */ - -void rfof_interface_set_marker_pointers(void** cptr_marker, int* id, +void rfof_set_marker_manually(void** cptr_marker, int* id, real* weight, real* R, real* phi, real* z, real* psi, real* charge, - real* mass, real* Ekin, real* velocity, real* mu, real* pphicanonical, - real* vpar, real* vperp, real* gyrof, real* vdriftRho, real* acc, - int* isOrbitTimeAccelerated, int* is_already_allocated); - - -/* KICKS */ - -void rfof_interface_do_rfof_stuff_gc(particle_simd_gc* ascot_marker, real* hin, - real* hout_rfof, rfof_data rfof_data, B_field_data* Bdata, - void** rfof_marker_pointer_array, void** rfof_mem_pointer_array, - void** rfof_diag_pointer_array, int* mem_shape_i, int* mem_shape_j); - - -/* RESET RESONANCE MEMORY */ - -void rfof_interface_reset_icrh_mem(void** rfof_marker_pointer, int* mem_shape_i, - int* mem_shape_j); - - -/* DEALLOCATION ROUTINES */ - -void rfof_interface_deallocate_rfof_input_param(void** cptr_rfof_input_param); - -void rfof_interface_deallocate_rfglobal(void** cptr_rfglobal); - -void rfof_interface_deallocate_res_mem(void** cptr_res_mem, - int* cptr_mem_shape_i, int* cptr_mem_shape_j); - -void rfof_interface_deallocate_diagnostics(void** cptr_diagno); + real* mass, real* Ekin, real* vnorm, real* mu, real* Pphi, + real* vpar, real* vperp, real* gyrof, real* vdriftRho, real* acceleration, + int* is_accelerated, int* is_already_allocated); -void rfof_interface_deallocate_marker(void** cptr_rfof_marker); +void rfof_set_up(rfof_marker* rfof_mrk, rfof_data* rfof_data); +void rfof_tear_down(rfof_marker* rfof_mrk); -/* FOR VISUALISING ICRH WAVE FIELD AND RESONANCE */ +void rfof_clear_history(rfof_marker* rfof_mrk, int imrk); -void rfof_interface_get_rf_wave_local(real* R, real* z, real* rho_tor, - real* theta, void** cptr_wi, real* e_plus_real, real* e_minus_real, - real* e_plus_imag, real* e_minus_imag); +void rfof_resonance_check_and_kick_gc( + particle_simd_gc* p, real* hin, real* hout_rfof, rfof_marker* rfof_mrk, + rfof_data* rfof_data, B_field_data* Bdata); -void rfof_interface_eval_resonance_function(void** cptr_marker, - void** cptr_rfglobal, real* omega_res, int* nharm); +void rfof_eval_rf_wave( + real* e_plus_real, real* e_minus_real, real* e_plus_imag, + real* e_minus_imag, real R, real z, rfof_data* rfof); +void rfof_eval_resonance_function( + real* omega_res, int* nharm, rfof_marker* rfof_mrk, rfof_data* rfof); #endif diff --git a/src/simulate.c b/src/simulate.c index 103f4200..777cc0cf 100644 --- a/src/simulate.c +++ b/src/simulate.c @@ -27,7 +27,7 @@ #include "simulate/mccc/mccc.h" #include "gctransform.h" #include "asigma.h" -#include "rfof_interface.h" +#include "rfof.h" #pragma omp declare target void sim_monitor(char* filename, volatile int* n, volatile int* finished); diff --git a/src/simulate.h b/src/simulate.h index 65fa6bbe..43de7b4e 100644 --- a/src/simulate.h +++ b/src/simulate.h @@ -22,7 +22,7 @@ #include "offload.h" #include "random.h" #include "simulate/mccc/mccc.h" -#include "rfof_interface.h" +#include "rfof.h" /** * @brief Simulaton modes diff --git a/src/simulate/simulate_gc_adaptive.c b/src/simulate/simulate_gc_adaptive.c index 41c41cd0..c8f681ce 100644 --- a/src/simulate/simulate_gc_adaptive.c +++ b/src/simulate/simulate_gc_adaptive.c @@ -20,7 +20,7 @@ #include "../E_field.h" #include "../boozer.h" #include "../mhd.h" -#include "../rfof_interface.h" +#include "../rfof.h" #include "../plasma.h" #include "simulate_gc_adaptive.h" #include "step/step_gc_cashkarp.h" @@ -61,11 +61,11 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { /* Current time step, suggestions for the next time step and next time * step */ - real hin[NSIMD] __memalign__; - real hout_orb[NSIMD] __memalign__; - real hout_col[NSIMD] __memalign__; + real hin[NSIMD] __memalign__; + real hout_orb[NSIMD] __memalign__; + real hout_col[NSIMD] __memalign__; real hout_rfof[NSIMD] __memalign__; - real hnext[NSIMD] __memalign__; + real hnext[NSIMD] __memalign__; /* Flag indicateing whether a new marker was initialized */ int cycle[NSIMD] __memalign__; @@ -78,6 +78,8 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { particle_simd_gc p; // This array holds current states particle_simd_gc p0; // This array stores previous states + rfof_marker rfof_mrk; // RFOF specific data + for(int i=0; i< NSIMD; i++) { p.id[i] = -1; p.running[i] = 0; @@ -86,41 +88,8 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { /* Initialize running particles */ int n_running = particle_cycle_gc(pq, &p, &sim->B_data, cycle); - /** @brief C equivalents of Fortran pointers to RFOF (ICRH) markers. */ - void* rfof_marker_pointer_array[NSIMD] __memalign__; - - /** @brief C equivalents of fortran pointers to resonance memorys of rfof - markers */ - void* rfof_mem_pointer_array[NSIMD] __memalign__; - - /** @brief C equivalents of fortran diagnostics pointers. These are not - * really used but they must be allocated nevertheless if one does - * not want a segmentation fault. */ - void* rfof_diag_pointer_array[NSIMD] __memalign__; - - /** @brief Number of rows in an RFOF resonance memory matrix. */ - int mem_shape_i[NSIMD] __memalign__; - - /** @brief Number of columns in an RFOF resonance memory matrix. */ - int mem_shape_j[NSIMD] __memalign__; - if(sim->enable_icrh) { - for(int i=0; i< NSIMD; i++) { - /* Allocate memory for the rfof markers on the fortran side. */ - rfof_interface_allocate_rfof_marker( - &(rfof_marker_pointer_array[i])); - - /* Allocate memory for the rfof resonance memory on the fortran - side. */ - rfof_interface_initialise_res_mem(&(rfof_mem_pointer_array[i]), - &(mem_shape_i[i]), &(mem_shape_j[i]), - &(sim->rfof_data.cptr_rfglobal), - &(sim->rfof_data.cptr_rfof_input_params)); - - /* Initialise rfof diagnostics (dummy argument for ICRH kick) */ - rfof_interface_initialise_diagnostics( - &(sim->rfof_data.cptr_rfglobal), &(rfof_diag_pointer_array[i])); - } + rfof_set_up(&rfof_mrk, &sim->rfof_data); } #pragma omp simd @@ -155,10 +124,10 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { #pragma omp simd for(int i = 0; i < NSIMD; i++) { particle_copy_gc(&p, i, &p0, i); - hout_orb[i] = DUMMY_TIMESTEP_VAL; - hout_col[i] = DUMMY_TIMESTEP_VAL; + hout_orb[i] = DUMMY_TIMESTEP_VAL; + hout_col[i] = DUMMY_TIMESTEP_VAL; hout_rfof[i] = DUMMY_TIMESTEP_VAL; - hnext[i] = DUMMY_TIMESTEP_VAL; + hnext[i] = DUMMY_TIMESTEP_VAL; } /*************************** Physics **********************************/ @@ -215,13 +184,10 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { } } - /* TODO implement RFOF_kick */ + /* Performs the ICRH kick if in resonance. */ if(sim->enable_icrh) { - /* Performs the ICRH kick if in resonance. */ - rfof_interface_do_rfof_stuff_gc(&p, hin, hout_rfof, sim->rfof_data, - &(sim->B_data), rfof_marker_pointer_array, - rfof_mem_pointer_array, rfof_diag_pointer_array, mem_shape_i, - mem_shape_j); + rfof_resonance_check_and_kick_gc( + &p, hin, hout_rfof, &rfof_mrk, &sim->rfof_data, &sim->B_data); /* Check whether time step was rejected */ #pragma omp simd @@ -324,8 +290,7 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { } if(sim->enable_icrh) { /* Reset icrh (rfof) resonance memory matrix. */ - rfof_interface_reset_icrh_mem(&(rfof_mem_pointer_array[i]), - &(mem_shape_i[i]), &(mem_shape_j[i])); + rfof_clear_history(&rfof_mrk, i); } } } @@ -333,22 +298,9 @@ void simulate_gc_adaptive(particle_queue* pq, sim_data* sim) { /* All markers simulated! */ - /* TODO: deallocate rfof structs (excl. wave field and input param which are - read in only once) */ /* Deallocate rfof structs */ if(sim->enable_icrh) { - for(int i=0; i< NSIMD; i++) { - /* Deallocate rfof markers */ - rfof_interface_deallocate_marker(&(rfof_marker_pointer_array[i])); - - /* Deallocate rfof marker resonance memory matrix. */ - rfof_interface_deallocate_res_mem(&(rfof_mem_pointer_array[i]), - &(mem_shape_i[i]), &(mem_shape_j[i])); - - /* Deallocate dummy diagnostics of rfof markers. */ - rfof_interface_deallocate_diagnostics( - &(rfof_diag_pointer_array[i])); - } + rfof_tear_down(&rfof_mrk); } }