api_cuda_cfd_coupling

2

5

6namespace rocky::cuda

7{

8

9template <class _Plugin= device::EmptyCustomCouplingPlugin>

10void select_and_dispatch_cfd_coupling_kernel(

11 IRockyModel& model,

12 DeviceParticleGroupData* device_group_data,

13 void** udf_devices_data

14)

15{

16 if (model.get_number_of_active_particles() <= 0)

17 return;

18

19 // It will be enough to have a data recorder

20 bool collect_cfd_data = model._impl->GetDeviceModel(0).cfd_data_recorder != nullptr;

21

22 for (int device_index = 0; device_index<model.get_number_of_devices(); device_index++)

23 {

24 void* udf_device_data = (udf_devices_data != nullptr ? udf_devices_data[device_index] : nullptr);

25 SDeviceModel* device_data = model.get_device_data(device_index);

26 const auto& group_device_data = device_group_data[device_index];

27

28 CUDA_SET_DEVICE(model.get_device_id(device_index));

29

30 unsigned int* group_particle_indexes = group_device_data.d_group_2_particles;

31 for (auto j = 0; j < group_device_data.group_indexes.size(); ++j)

32 {

33 const int particle_group_index = group_device_data.group_indexes[j];

34 device::select_and_dispatch_coupling_kernel<_Plugin>(

35 true,

36 collect_cfd_data,

37 device_data,

38 group_particle_indexes,

39 group_device_data.n_particles_per_group[j],

40 udf_device_data,

41 model._impl->cfd_data.gen_cfd,

42 model._impl->cfd_data.cfd_particle_groups[particle_group_index]);

43 group_particle_indexes += group_device_data.n_particles_per_group[j];

44 }

45 }

46

47 //Sync all GPUs

48 sync_devices(model);

49}

50

51template <typename _Op>

55__global__ void cuda_dispatch_fluid_kernel(SDeviceModel* device_data, double current_time, _Op op, void* udf_data)

56{

57 device::api_compute_cfd_data_coupling<CudaBackend>(device_data, current_time, op, udf_data);

58}

59

60template <typename _Op>

62void dispatch_fluid_kernel(IRockyModel& model, _Op op, void** udf_devices_data)

63{

64 if (model.get_number_of_active_particles() <= 0)

65 return;

66

67 for (int i = 0; i<model.get_number_of_devices(); i++)

68 {

69 CUDA_SET_DEVICE(model.get_device_id(i));

70

71 cuda_dispatch_fluid_kernel<<<CD_NBLOCKS, CD_NTHREADS>>>(

72 model.get_device_data(i),

73 model.get_current_time(),

74 op,

75 (udf_devices_data != nullptr ? udf_devices_data[i] : nullptr)

76 );

77 }

78 //Sync all GPUs

79 sync_devices(model);

80}

81

82}