## On this page

  

 

 # api\_cuda\_cfd\_coupling

 Last update: 17.07.2025 

<a id="l00001" name="l00001"></a> 1\#pragma once

<a id="l00002" name="l00002"></a> 2 

<a id="l00003" name="l00003"></a> 3\#include &lt;rocky20/api/device/api\_cuda.hpp&gt;

<a id="l00004" name="l00004"></a> 4\#include &lt;rocky20/api/device/api\_device\_cfd\_coupling.hpp&gt;

<a id="l00005" name="l00005"></a> 5 

<a id="l00006" name="l00006"></a> 6namespace rocky::cuda

<a id="l00007" name="l00007"></a> 7{

<a id="l00008" name="l00008"></a> 8 

<a id="l00009" name="l00009"></a> 9template &lt;class _Plugin= device::EmptyCustomCouplingPlugin&gt;

<a id="l00010" name="l00010"></a> 10void select_and_dispatch_cfd_coupling_kernel(

<a id="l00011" name="l00011"></a> 11 [IRockyModel](structIRockyModel.xhtml)&amp; model,

<a id="l00012" name="l00012"></a> 12 DeviceParticleGroupData* device_group_data,

<a id="l00013" name="l00013"></a> 13 void** udf_devices_data

<a id="l00014" name="l00014"></a> 14)

<a id="l00015" name="l00015"></a> 15{

<a id="l00016" name="l00016"></a> 16 if (model.[get\_number\_of\_active\_particles](structIRockyModel.xhtml#a38a15145f4f2bb47b79ab785f50d8277)() &lt;= 0)

<a id="l00017" name="l00017"></a> 17 return;

<a id="l00018" name="l00018"></a> 18 

<a id="l00019" name="l00019"></a> 19 // It will be enough to have a data recorder

<a id="l00020" name="l00020"></a> 20 bool collect_cfd_data = model._impl-&gt;GetDeviceModel(0).cfd_data_recorder != nullptr;

<a id="l00021" name="l00021"></a> 21 

<a id="l00022" name="l00022"></a> 22 for (int device_index = 0; device_index&lt;model.[get\_number\_of\_devices](structIRockyModel.xhtml#a1fd1011aa809fe64647e9cacf1956a11)(); device_index++)

<a id="l00023" name="l00023"></a> 23 {

<a id="l00024" name="l00024"></a> 24 void* udf_device_data = (udf_devices_data != nullptr ? udf_devices_data[device_index] : nullptr);

<a id="l00025" name="l00025"></a> 25 SDeviceModel* device_data = model.get_device_data(device_index);

<a id="l00026" name="l00026"></a> 26 const auto&amp; group_device_data = device_group_data[device_index];

<a id="l00027" name="l00027"></a> 27 

<a id="l00028" name="l00028"></a> 28 CUDA_SET_DEVICE(model.[get\_device\_id](structIRockyModel.xhtml#ae8f8865c809a31988291f9180d7cda2b)(device_index));

<a id="l00029" name="l00029"></a> 29 

<a id="l00030" name="l00030"></a> 30 unsigned int* group_particle_indexes = group_device_data.d_group_2_particles;

<a id="l00031" name="l00031"></a> 31 for (auto j = 0; j &lt; group_device_data.group_indexes.size(); ++j)

<a id="l00032" name="l00032"></a> 32 {

<a id="l00033" name="l00033"></a> 33 const int particle_group_index = group_device_data.group_indexes[j];

<a id="l00034" name="l00034"></a> 34 device::select_and_dispatch_coupling_kernel&lt;_Plugin&gt;(

<a id="l00035" name="l00035"></a> 35 true,

<a id="l00036" name="l00036"></a> 36 collect_cfd_data,

<a id="l00037" name="l00037"></a> 37 device_data,

<a id="l00038" name="l00038"></a> 38 group_particle_indexes,

<a id="l00039" name="l00039"></a> 39 group_device_data.n_particles_per_group[j],

<a id="l00040" name="l00040"></a> 40 udf_device_data,

<a id="l00041" name="l00041"></a> 41 model._impl-&gt;cfd_data.gen_cfd,

<a id="l00042" name="l00042"></a> 42 model._impl-&gt;cfd_data.cfd_particle_groups[particle_group_index]);

<a id="l00043" name="l00043"></a> 43 group_particle_indexes += group_device_data.n_particles_per_group[j];

<a id="l00044" name="l00044"></a> 44 }

<a id="l00045" name="l00045"></a> 45 }

<a id="l00046" name="l00046"></a> 46 

<a id="l00047" name="l00047"></a> 47 //Sync all GPUs

<a id="l00048" name="l00048"></a> 48 sync_devices(model);

<a id="l00049" name="l00049"></a> 49}

<a id="l00050" name="l00050"></a> 50 

<a id="l00051" name="l00051"></a> 51template &lt;typename _Op&gt;

<a id="l00052" name="l00052"></a> 52\#ifndef \_\_CUDACC\_\_

<a id="l00053" name="l00053"></a> 53inline

<a id="l00054" name="l00054"></a> 54\#endif

<a id="l00055" name="l00055"></a> 55__global__ void cuda_dispatch_fluid_kernel(SDeviceModel* device_data, double current_time, _Op op, void* udf_data)

<a id="l00056" name="l00056"></a> 56{

<a id="l00057" name="l00057"></a> 57 device::api_compute_cfd_data_coupling&lt;CudaBackend&gt;(device_data, current_time, op, udf_data);

<a id="l00058" name="l00058"></a> 58}

<a id="l00059" name="l00059"></a> 59 

<a id="l00060" name="l00060"></a> 60template &lt;typename _Op&gt;

<a id="l00061" name="l00061"></a> 61inline

<a id="l00062" name="l00062"></a> 62void dispatch_fluid_kernel([IRockyModel](structIRockyModel.xhtml)&amp; model, _Op op, void** udf_devices_data)

<a id="l00063" name="l00063"></a> 63{

<a id="l00064" name="l00064"></a> 64 if (model.[get\_number\_of\_active\_particles](structIRockyModel.xhtml#a38a15145f4f2bb47b79ab785f50d8277)() &lt;= 0)

<a id="l00065" name="l00065"></a> 65 return;

<a id="l00066" name="l00066"></a> 66 

<a id="l00067" name="l00067"></a> 67 for (int i = 0; i&lt;model.[get\_number\_of\_devices](structIRockyModel.xhtml#a1fd1011aa809fe64647e9cacf1956a11)(); i++)

<a id="l00068" name="l00068"></a> 68 {

<a id="l00069" name="l00069"></a> 69 CUDA_SET_DEVICE(model.[get\_device\_id](structIRockyModel.xhtml#ae8f8865c809a31988291f9180d7cda2b)(i));

<a id="l00070" name="l00070"></a> 70 

<a id="l00071" name="l00071"></a> 71 cuda_dispatch_fluid_kernel&lt;&lt;&lt;CD_NBLOCKS, CD_NTHREADS&gt;&gt;&gt;(

<a id="l00072" name="l00072"></a> 72 model.get_device_data(i),

<a id="l00073" name="l00073"></a> 73 model.[get\_current\_time](structIRockyModel.xhtml#a93ef35ae1eb9cf50ca77ab2070a769ab)(),

<a id="l00074" name="l00074"></a> 74 op,

<a id="l00075" name="l00075"></a> 75 (udf_devices_data != nullptr ? udf_devices_data[i] : nullptr)

<a id="l00076" name="l00076"></a> 76 );

<a id="l00077" name="l00077"></a> 77 }

<a id="l00078" name="l00078"></a> 78 //Sync all GPUs

<a id="l00079" name="l00079"></a> 79 sync_devices(model);

<a id="l00080" name="l00080"></a> 80}

<a id="l00081" name="l00081"></a> 81 

<a id="l00082" name="l00082"></a> 82}

[IRockyModel](structIRockyModel.xhtml)

**Definition** rocky_api.h:60



[IRockyModel::get\_number\_of\_devices](structIRockyModel.xhtml#a1fd1011aa809fe64647e9cacf1956a11)

int get_number_of_devices() const

**Definition** rocky_api.cpp:159



[IRockyModel::get\_number\_of\_active\_particles](structIRockyModel.xhtml#a38a15145f4f2bb47b79ab785f50d8277)

int get_number_of_active_particles() const

**Definition** rocky_api.cpp:139



[IRockyModel::get\_current\_time](structIRockyModel.xhtml#a93ef35ae1eb9cf50ca77ab2070a769ab)

double get_current_time() const

**Definition** rocky_api.cpp:129



[IRockyModel::get\_device\_id](structIRockyModel.xhtml#ae8f8865c809a31988291f9180d7cda2b)

int get_device_id(int device_index) const

**Definition** rocky_api.cpp:164