## On this page

  

 

 # api\_cuda\_joints

 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\_device\_joints.hpp&gt;

<a id="l00004" name="l00004"></a> 4\#include &lt;rocky20/api/device/api\_cuda.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;typename Callable&gt;

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

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

<a id="l00012" name="l00012"></a> 12 Callable h_joint_f_ptr,

<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 for (int i = 0; i &lt; model.[get\_number\_of\_devices](structIRockyModel.xhtml#a1fd1011aa809fe64647e9cacf1956a11)(); i++)

<a id="l00017" name="l00017"></a> 17 {

<a id="l00018" name="l00018"></a> 18 void *udf_device_data = udf_devices_data[i];

<a id="l00019" name="l00019"></a> 19 SDeviceModel *device_data = model.get_device_data(i);

<a id="l00020" name="l00020"></a> 20 

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

<a id="l00022" name="l00022"></a> 22 device::execute_joints_function&lt;CudaBackend&gt;&lt;&lt;&lt;CD_NBLOCKS, CD_NTHREADS&gt;&gt;&gt;(

<a id="l00023" name="l00023"></a> 23 device_data, i, model.[get\_current\_time](structIRockyModel.xhtml#a93ef35ae1eb9cf50ca77ab2070a769ab)(),

<a id="l00024" name="l00024"></a> 24 h_joint_f_ptr,

<a id="l00025" name="l00025"></a> 25 udf_device_data

<a id="l00026" name="l00026"></a> 26 );

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

<a id="l00028" name="l00028"></a> 28 

<a id="l00029" name="l00029"></a> 29 //Sync all GPUs

<a id="l00030" name="l00030"></a> 30 sync_devices(model);

<a id="l00031" name="l00031"></a> 31}

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

<a id="l00033" name="l00033"></a> 33template &lt;typename Callable&gt;

<a id="l00034" name="l00034"></a> 34__global__ void execute_joint_forces_function_kernel(

<a id="l00035" name="l00035"></a> 35 SDeviceModel *device_data, int device_id,

<a id="l00036" name="l00036"></a> 36 Callable joint_f_ptr,

<a id="l00037" name="l00037"></a> 37 void *udf_device_data

<a id="l00038" name="l00038"></a> 38)

<a id="l00039" name="l00039"></a> 39{

<a id="l00040" name="l00040"></a> 40 device::compute_joint_forces&lt;CudaBackend&gt;(

<a id="l00041" name="l00041"></a> 41 device_data, device_id,

<a id="l00042" name="l00042"></a> 42 joint_f_ptr, udf_device_data

<a id="l00043" name="l00043"></a> 43 );

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

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

<a id="l00046" name="l00046"></a> 46template &lt;typename Callable&gt;

<a id="l00047" name="l00047"></a> 47__global__ void execute_joint_breakage_function_kernel(

<a id="l00048" name="l00048"></a> 48 SDeviceModel* device_data, int device_id,

<a id="l00049" name="l00049"></a> 49 Callable joint_breakage_f_ptr,

<a id="l00050" name="l00050"></a> 50 void* udf_device_data,

<a id="l00051" name="l00051"></a> 51 typename CudaBackend::atomic_int_ptr nbreak_joints_a,

<a id="l00052" name="l00052"></a> 52 cu_real time_curr

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

<a id="l00054" name="l00054"></a> 54{

<a id="l00055" name="l00055"></a> 55 device::compute_joint_breakage&lt;CudaBackend&gt;(

<a id="l00056" name="l00056"></a> 56 device_data, device_id,

<a id="l00057" name="l00057"></a> 57 joint_breakage_f_ptr, udf_device_data,

<a id="l00058" name="l00058"></a> 58 nbreak_joints_a,

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

<a id="l00060" name="l00060"></a> 60 );

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

<a id="l00062" name="l00062"></a> 62 

<a id="l00063" name="l00063"></a> 63template &lt;typename Callable&gt;

<a id="l00064" name="l00064"></a> 64void dispatch_joint_forces_kernel(

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

<a id="l00066" name="l00066"></a> 66 Callable h_joint_f_ptr,

<a id="l00067" name="l00067"></a> 67 void **udf_devices_data

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

<a id="l00069" name="l00069"></a> 69{

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

<a id="l00071" name="l00071"></a> 71 {

<a id="l00072" name="l00072"></a> 72 void *udf_device_data = udf_devices_data[i];

<a id="l00073" name="l00073"></a> 73 SDeviceModel *device_data = model.get_device_data(i);

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

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

<a id="l00076" name="l00076"></a> 76 rocky::cuda::execute_joint_forces_function_kernel &lt;&lt; &lt;CD_NBLOCKS, CD_NTHREADS &gt;&gt; &gt; (

<a id="l00077" name="l00077"></a> 77 device_data, i,

<a id="l00078" name="l00078"></a> 78 h_joint_f_ptr,

<a id="l00079" name="l00079"></a> 79 udf_device_data

<a id="l00080" name="l00080"></a> 80 );

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

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

<a id="l00083" name="l00083"></a> 83 //Sync all GPUs

<a id="l00084" name="l00084"></a> 84 sync_devices(model);

<a id="l00085" name="l00085"></a> 85}

<a id="l00086" name="l00086"></a> 86 

<a id="l00087" name="l00087"></a> 87template &lt;typename Callable&gt;

<a id="l00088" name="l00088"></a> 88void dispatch_joint_breakage_kernel(

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

<a id="l00090" name="l00090"></a> 90 Callable h_joint_breakage_f_ptr,

<a id="l00091" name="l00091"></a> 91 void** udf_devices_data

<a id="l00092" name="l00092"></a> 92)

<a id="l00093" name="l00093"></a> 93{ 

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

<a id="l00095" name="l00095"></a> 95 {

<a id="l00096" name="l00096"></a> 96 void* udf_device_data = udf_devices_data[i];

<a id="l00097" name="l00097"></a> 97 SDeviceModel* device_data = model.get_device_data(i); 

<a id="l00098" name="l00098"></a> 98 

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

<a id="l00100" name="l00100"></a> 100 rocky::cuda::execute_joint_breakage_function_kernel &lt;&lt;&lt;CD_NBLOCKS, CD_NTHREADS &gt;&gt;&gt; (

<a id="l00101" name="l00101"></a> 101 device_data, i,

<a id="l00102" name="l00102"></a> 102 h_joint_breakage_f_ptr,

<a id="l00103" name="l00103"></a> 103 udf_device_data,

<a id="l00104" name="l00104"></a> 104 &amp;(model._impl-&gt;device_models[i].gen-&gt;nbreak_joints),

<a id="l00105" name="l00105"></a> 105 model._impl-&gt;iteration.current_time

<a id="l00106" name="l00106"></a> 106 );

<a id="l00107" name="l00107"></a> 107 }

<a id="l00108" name="l00108"></a> 108 

<a id="l00109" name="l00109"></a> 109 //Sync all GPUs

<a id="l00110" name="l00110"></a> 110 sync_devices(model);

<a id="l00111" name="l00111"></a> 111}

<a id="l00112" name="l00112"></a> 112 

<a id="l00113" name="l00113"></a> 113} // namespace rocky::cuda

[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\_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