310_data_management.doxy 35 KB

123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566676869707172737475767778798081828384858687888990919293949596979899100101102103104105106107108109110111112113114115116117118119120121122123124125126127128129130131132133134135136137138139140141142143144145146147148149150151152153154155156157158159160161162163164165166167168169170171172173174175176177178179180181182183184185186187188189190191192193194195196197198199200201202203204205206207208209210211212213214215216217218219220221222223224225226227228229230231232233234235236237238239240241242243244245246247248249250251252253254255256257258259260261262263264265266267268269270271272273274275276277278279280281282283284285286287288289290291292293294295296297298299300301302303304305306307308309310311312313314315316317318319320321322323324325326327328329330331332333334335336337338339340341342343344345346347348349350351352353354355356357358359360361362363364365366367368369370371372373374375376377378379380381382383384385386387388389390391392393394395396397398399400401402403404405406407408409410411412413414415416417418419420421422423424425426427428429430431432433434435436437438439440441442443444445446447448449450451452453454455456457458459460461462463464465466467468469470471472473474475476477478479480481482483484485486487488489490491492493494495496497498499500501502503504505506507508509510511512513514515516517518519520521522523524525526527528529530531532533534535536537538539540541542543544545546547548549550551552553554555556557558559560561562563564565566567568569570571572573574575576577578579580581582583584585586587588589590591592593594595596597598599600601602603604605606607608609610611612613614615616617618619620621622623624625626627628629630631632633634635636637638639640641642643644645646647648649650651652653654655656657658659660661662663664665666667668669670671672673674675676677678679680681682683684685686687688689690691692693694695696697698699700701702703704705706707708709710711712713714715716717718719720721722723724725726727728729730731732733734735736737738739740741742743744745746747748749750751752753754755756757758759760761762763764765766767768769770771772773774775776777778779780781782783784785786787788789790791792793794795796797798799800801802803804805806807808809810811812813814815816817818819820821822823824825826827828829830831832833834835836837838839840841842843844845846847848849850851852853854855856857858859860861862863864865866867868869870871872873874875876877878879880881882883884885886887888889890891892893894895896897898899900901
  1. /* StarPU --- Runtime system for heterogeneous multicore architectures.
  2. *
  3. * Copyright (C) 2010-2019 CNRS
  4. * Copyright (C) 2009-2011,2014-2019 Université de Bordeaux
  5. * Copyright (C) 2011,2012 Inria
  6. *
  7. * StarPU is free software; you can redistribute it and/or modify
  8. * it under the terms of the GNU Lesser General Public License as published by
  9. * the Free Software Foundation; either version 2.1 of the License, or (at
  10. * your option) any later version.
  11. *
  12. * StarPU is distributed in the hope that it will be useful, but
  13. * WITHOUT ANY WARRANTY; without even the implied warranty of
  14. * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.
  15. *
  16. * See the GNU Lesser General Public License in COPYING.LGPL for more details.
  17. */
  18. /*! \page DataManagement Data Management
  19. TODO: intro which mentions consistency among other things
  20. \section DataInterface Data Interface
  21. StarPU provides several data interfaces for programmers to describe the data layout of their application. There are predefined interfaces already available in StarPU. Users can define new data interfaces as explained in \ref DefiningANewDataInterface. All functions provided by StarPU are documented in \ref API_Data_Interfaces. You will find a short list below.
  22. \subsection VariableDataInterface Variable Data Interface
  23. A variable is a given size byte element, typically a scalar. Here an
  24. example of how to register a variable data to StarPU by using
  25. starpu_variable_data_register().
  26. \code{.c}
  27. float var = 42.0;
  28. starpu_data_handle_t var_handle;
  29. starpu_variable_data_register(&var_handle, STARPU_MAIN_RAM, (uintptr_t)&var, sizeof(var));
  30. \endcode
  31. \subsection VectorDataInterface Vector Data Interface
  32. A vector is a fixed number of elements of a given size. Here an
  33. example of how to register a vector data to StarPU by using
  34. starpu_vector_data_register().
  35. \code{.c}
  36. float vector[NX];
  37. starpu_data_handle_t vector_handle;
  38. starpu_vector_data_register(&vector_handle, STARPU_MAIN_RAM, (uintptr_t)vector, NX, sizeof(vector[0]));
  39. \endcode
  40. \subsection MatrixDataInterface Matrix Data Interface
  41. To register 2-D matrices with a potential padding, one can use the
  42. matrix data interface. Here an example of how to register a matrix
  43. data to StarPU by using starpu_matrix_data_register().
  44. \code{.c}
  45. float *matrix;
  46. starpu_data_handle_t matrix_handle;
  47. matrix = (float*)malloc(width * height * sizeof(float));
  48. starpu_matrix_data_register(&matrix_handle, STARPU_MAIN_RAM, (uintptr_t)matrix, width, width, height, sizeof(float));
  49. \endcode
  50. \subsection BlockDataInterface Block Data Interface
  51. To register 3-D blocks with potential paddings on Y and Z dimensions,
  52. one can use the block data interface. Here an example of how to
  53. register a block data to StarPU by using starpu_block_data_register().
  54. \code{.c}
  55. float *block;
  56. starpu_data_handle_t block_handle;
  57. block = (float*)malloc(nx*ny*nz*sizeof(float));
  58. starpu_block_data_register(&block_handle, STARPU_MAIN_RAM, (uintptr_t)block, nx, nx*ny, nx, ny, nz, sizeof(float));
  59. \endcode
  60. \subsection BCSRDataInterface BCSR Data Interface
  61. BCSR (Blocked Compressed Sparse Row Representation) sparse matrix data
  62. can be registered to StarPU using the bcsr data interface. Here an
  63. example on how to do so by using starpu_bcsr_data_register().
  64. \code{.c}
  65. /*
  66. * We use the following matrix:
  67. *
  68. * +----------------+
  69. * | 0 1 0 0 |
  70. * | 2 3 0 0 |
  71. * | 4 5 8 9 |
  72. * | 6 7 10 11 |
  73. * +----------------+
  74. *
  75. * nzval = [0, 1, 2, 3] ++ [4, 5, 6, 7] ++ [8, 9, 10, 11]
  76. * colind = [0, 0, 1]
  77. * rowptr = [0, 1, 3]
  78. * r = c = 2
  79. */
  80. /* Size of the blocks */
  81. int R = 2;
  82. int C = 2;
  83. int NROWS = 2;
  84. int NNZ_BLOCKS = 3; /* out of 4 */
  85. int NZVAL_SIZE = (R*C*NNZ_BLOCKS);
  86. int nzval[NZVAL_SIZE] =
  87. {
  88. 0, 1, 2, 3, /* First block */
  89. 4, 5, 6, 7, /* Second block */
  90. 8, 9, 10, 11 /* Third block */
  91. };
  92. uint32_t colind[NNZ_BLOCKS] =
  93. {
  94. 0, /* block-column index for first block in nzval */
  95. 0, /* block-column index for second block in nzval */
  96. 1 /* block-column index for third block in nzval */
  97. };
  98. uint32_t rowptr[NROWS+1] =
  99. {
  100. 0, / * block-index in nzval of the first block of the first row. */
  101. 1, / * block-index in nzval of the first block of the second row. */
  102. NNZ_BLOCKS /* number of blocks, to allow an easier element's access for the kernels */
  103. };
  104. starpu_data_handle_t bcsr_handle;
  105. starpu_bcsr_data_register(&bcsr_handle,
  106. STARPU_MAIN_RAM,
  107. NNZ_BLOCKS,
  108. NROWS,
  109. (uintptr_t) nzval,
  110. colind,
  111. rowptr,
  112. 0, /* firstentry */
  113. R,
  114. C,
  115. sizeof(nzval[0]));
  116. \endcode
  117. StarPU provides an example on how to deal with such matrices in
  118. <c>examples/spmv</c>.
  119. \subsection CSRDataInterface CSR Data Interface
  120. TODO
  121. \subsection VariableSizeDataInterface Data Interface with Variable Size
  122. Tasks are actually allowed to change the size of data interfaces.
  123. The task implementation can just reallocate the buffer during its execution, and
  124. set the proper new values in the interface structure, e.g. nx, ny, ld, etc. so
  125. that the StarPU core knows the new data layout. The starpu_data_interface_ops
  126. structure however then needs to have the starpu_data_interface_ops::dontcache
  127. field set to 1, to prevent StarPU from trying to perform any cached allocation,
  128. since the allocated size will vary. An example is available in
  129. <c>tests/datawizard/variable_size.c</c>
  130. \section DataManagement Data Management
  131. When the application allocates data, whenever possible it should use
  132. the starpu_malloc() function, which will ask CUDA or OpenCL to make
  133. the allocation itself and pin the corresponding allocated memory, or to use the
  134. starpu_memory_pin() function to pin memory allocated by other ways, such as local arrays. This
  135. is needed to permit asynchronous data transfer, i.e. permit data
  136. transfer to overlap with computations. Otherwise, the trace will show
  137. that the <c>DriverCopyAsync</c> state takes a lot of time, this is
  138. because CUDA or OpenCL then reverts to synchronous transfers.
  139. By default, StarPU leaves replicates of data wherever they were used, in case they
  140. will be re-used by other tasks, thus saving the data transfer time. When some
  141. task modifies some data, all the other replicates are invalidated, and only the
  142. processing unit which ran this task will have a valid replicate of the data. If the application knows
  143. that this data will not be re-used by further tasks, it should advise StarPU to
  144. immediately replicate it to a desired list of memory nodes (given through a
  145. bitmask). This can be understood like the write-through mode of CPU caches.
  146. \code{.c}
  147. starpu_data_set_wt_mask(img_handle, 1<<0);
  148. \endcode
  149. will for instance request to always automatically transfer a replicate into the
  150. main memory (node <c>0</c>), as bit <c>0</c> of the write-through bitmask is being set.
  151. \code{.c}
  152. starpu_data_set_wt_mask(img_handle, ~0U);
  153. \endcode
  154. will request to always automatically broadcast the updated data to all memory
  155. nodes.
  156. Setting the write-through mask to <c>~0U</c> can also be useful to make sure all
  157. memory nodes always have a copy of the data, so that it is never evicted when
  158. memory gets scarse.
  159. Implicit data dependency computation can become expensive if a lot
  160. of tasks access the same piece of data. If no dependency is required
  161. on some piece of data (e.g. because it is only accessed in read-only
  162. mode, or because write accesses are actually commutative), use the
  163. function starpu_data_set_sequential_consistency_flag() to disable
  164. implicit dependencies on this data.
  165. In the same vein, accumulation of results in the same data can become a
  166. bottleneck. The use of the mode ::STARPU_REDUX permits to optimize such
  167. accumulation (see \ref DataReduction). To a lesser extent, the use of
  168. the flag ::STARPU_COMMUTE keeps the bottleneck (see \ref DataCommute), but at least permits
  169. the accumulation to happen in any order.
  170. Applications often need a data just for temporary results. In such a case,
  171. registration can be made without an initial value, for instance this produces a vector data:
  172. \code{.c}
  173. starpu_vector_data_register(&handle, -1, 0, n, sizeof(float));
  174. \endcode
  175. StarPU will then allocate the actual buffer only when it is actually needed,
  176. e.g. directly on the GPU without allocating in main memory.
  177. In the same vein, once the temporary results are not useful any more, the
  178. data should be thrown away. If the handle is not to be reused, it can be
  179. unregistered:
  180. \code{.c}
  181. starpu_data_unregister_submit(handle);
  182. \endcode
  183. actual unregistration will be done after all tasks working on the handle
  184. terminate.
  185. If the handle is to be reused, instead of unregistering it, it can simply be invalidated:
  186. \code{.c}
  187. starpu_data_invalidate_submit(handle);
  188. \endcode
  189. the buffers containing the current value will then be freed, and reallocated
  190. only when another task writes some value to the handle.
  191. \section DataPrefetch Data Prefetch
  192. The scheduling policies <c>heft</c>, <c>dmda</c> and <c>pheft</c>
  193. perform data prefetch (see \ref STARPU_PREFETCH):
  194. as soon as a scheduling decision is taken for a task, requests are issued to
  195. transfer its required data to the target processing unit, if needed, so that
  196. when the processing unit actually starts the task, its data will hopefully be
  197. already available and it will not have to wait for the transfer to finish.
  198. The application may want to perform some manual prefetching, for several reasons
  199. such as excluding initial data transfers from performance measurements, or
  200. setting up an initial statically-computed data distribution on the machine
  201. before submitting tasks, which will thus guide StarPU toward an initial task
  202. distribution (since StarPU will try to avoid further transfers).
  203. This can be achieved by giving the function starpu_data_prefetch_on_node() the
  204. handle and the desired target memory node. The
  205. starpu_data_idle_prefetch_on_node() variant can be used to issue the transfer
  206. only when the bus is idle.
  207. Conversely, one can advise StarPU that some data will not be useful in the
  208. close future by calling starpu_data_wont_use(). StarPU will then write its value
  209. back to its home node, and evict it from GPUs when room is needed.
  210. \section PartitioningData Partitioning Data
  211. An existing piece of data can be partitioned in sub parts to be used by different tasks, for instance:
  212. \code{.c}
  213. #define NX 1048576
  214. #define PARTS 16
  215. int vector[NX];
  216. starpu_data_handle_t handle;
  217. /* Declare data to StarPU */
  218. starpu_vector_data_register(&handle, STARPU_MAIN_RAM, (uintptr_t)vector, NX, sizeof(vector[0]));
  219. /* Partition the vector in PARTS sub-vectors */
  220. struct starpu_data_filter f =
  221. {
  222. .filter_func = starpu_vector_filter_block,
  223. .nchildren = PARTS
  224. };
  225. starpu_data_partition(handle, &f);
  226. \endcode
  227. The task submission then uses the function starpu_data_get_sub_data()
  228. to retrieve the sub-handles to be passed as tasks parameters.
  229. \code{.c}
  230. /* Submit a task on each sub-vector */
  231. for (i=0; i<starpu_data_get_nb_children(handle); i++)
  232. {
  233. /* Get subdata number i (there is only 1 dimension) */
  234. starpu_data_handle_t sub_handle = starpu_data_get_sub_data(handle, 1, i);
  235. struct starpu_task *task = starpu_task_create();
  236. task->handles[0] = sub_handle;
  237. task->cl = &cl;
  238. task->synchronous = 1;
  239. task->cl_arg = &factor;
  240. task->cl_arg_size = sizeof(factor);
  241. starpu_task_submit(task);
  242. }
  243. \endcode
  244. Partitioning can be applied several times, see
  245. <c>examples/basic_examples/mult.c</c> and <c>examples/filters/</c>.
  246. Wherever the whole piece of data is already available, the partitioning will
  247. be done in-place, i.e. without allocating new buffers but just using pointers
  248. inside the existing copy. This is particularly important to be aware of when
  249. using OpenCL, where the kernel parameters are not pointers, but \c cl_mem handles. The
  250. kernel thus needs to be also passed the offset within the OpenCL buffer:
  251. \code{.c}
  252. void opencl_func(void *buffers[], void *cl_arg)
  253. {
  254. cl_mem vector = (cl_mem) STARPU_VECTOR_GET_DEV_HANDLE(buffers[0]);
  255. unsigned offset = STARPU_BLOCK_GET_OFFSET(buffers[0]);
  256. ...
  257. clSetKernelArg(kernel, 0, sizeof(vector), &vector);
  258. clSetKernelArg(kernel, 1, sizeof(offset), &offset);
  259. ...
  260. }
  261. \endcode
  262. And the kernel has to shift from the pointer passed by the OpenCL driver:
  263. \code{.c}
  264. __kernel void opencl_kernel(__global int *vector, unsigned offset)
  265. {
  266. block = (__global void *)block + offset;
  267. ...
  268. }
  269. \endcode
  270. StarPU provides various interfaces and filters for matrices, vectors, etc.,
  271. but applications can also write their own data interfaces and filters, see
  272. <c>examples/interface</c> and <c>examples/filters/custom_mf</c> for an example,
  273. and see \ref DefiningANewDataInterface and \ref DefiningANewDataFilter
  274. for documentation.
  275. \section AsynchronousPartitioning Asynchronous Partitioning
  276. The partitioning functions described in the previous section are synchronous:
  277. starpu_data_partition() and starpu_data_unpartition() both wait for all the tasks
  278. currently working on the data. This can be a bottleneck for the application.
  279. An asynchronous API also exists, it works only on handles with sequential
  280. consistency. The principle is to first plan the partitioning, which returns
  281. data handles of the partition, which are not functional yet. When submitting
  282. tasks, one can mix using the handles of the partition, of the whole data. One
  283. can even partition recursively and mix using handles at different levels of the
  284. recursion. Of course, StarPU will have to introduce coherency synchronization.
  285. <c>fmultiple_submit_implicit</c> is a complete example using this technique.
  286. One can also look at <c>fmultiple_submit_readonly</c> which contains the
  287. explicit coherency synchronization which are automatically introduced by StarPU
  288. for <c>fmultiple_submit_implicit</c>.
  289. In short, we first register a matrix and plan the partitioning:
  290. \code{.c}
  291. starpu_matrix_data_register(&handle, STARPU_MAIN_RAM, (uintptr_t)matrix, NX, NX, NY, sizeof(matrix[0]));
  292. struct starpu_data_filter f_vert =
  293. {
  294. .filter_func = starpu_matrix_filter_block,
  295. .nchildren = PARTS
  296. };
  297. starpu_data_partition_plan(handle, &f_vert, vert_handle);
  298. \endcode
  299. starpu_data_partition_plan() returns the handles for the partition in <c>vert_handle</c>.
  300. One can then submit tasks working on the main handle, and tasks working on
  301. <c>vert_handle</c> handles. Between using the main handle and <c>vert_handle</c>
  302. handles, StarPU will automatically call starpu_data_partition_submit() and
  303. starpu_data_unpartition_submit().
  304. All this code is asynchronous, just submitting which tasks, partitioning and
  305. unpartitioning will be done at runtime.
  306. Planning several partitioning of the same data is also possible, StarPU will
  307. unpartition and repartition as needed when mixing accesses of different
  308. partitions. If data access is done in read-only mode, StarPU will allow the
  309. different partitioning to coexist. As soon as a data is accessed in read-write
  310. mode, StarPU will automatically unpartition everything and activate only the
  311. partitioning leading to the data being written to.
  312. For instance, for a stencil application, one can split a subdomain into
  313. its interior and halos, and then just submit a task updating the whole
  314. subdomain, then submit MPI sends/receives to update the halos, then submit
  315. again a task updating the whole subdomain, etc. and StarPU will automatically
  316. partition/unpartition each time.
  317. \section ManualPartitioning Manual Partitioning
  318. One can also handle partitioning by hand, by registering several views on the
  319. same piece of data. The idea is then to manage the coherency of the various
  320. views through the common buffer in the main memory.
  321. <c>fmultiple_manual</c> is a complete example using this technique.
  322. In short, we first register the same matrix several times:
  323. \code{.c}
  324. starpu_matrix_data_register(&handle, STARPU_MAIN_RAM, (uintptr_t)matrix, NX, NX, NY, sizeof(matrix[0]));
  325. for (i = 0; i < PARTS; i++)
  326. starpu_matrix_data_register(&vert_handle[i], STARPU_MAIN_RAM, (uintptr_t)&matrix[0][i*(NX/PARTS)], NX, NX/PARTS, NY, sizeof(matrix[0][0]));
  327. \endcode
  328. Since StarPU is not aware that the two handles are actually pointing to the same
  329. data, we have a danger of inadvertently submitting tasks to both views, which
  330. will bring a mess since StarPU will not guarantee any coherency between the two
  331. views. To make sure we don't do this, we invalidate the view that we will not
  332. use:
  333. \code{.c}
  334. for (i = 0; i < PARTS; i++)
  335. starpu_data_invalidate(vert_handle[i]);
  336. \endcode
  337. Then we can safely work on <c>handle</c>.
  338. When we want to switch to the vertical slice view, all we need to do is bring
  339. coherency between them by running an empty task on the home node of the data:
  340. \code{.c}
  341. void empty(void *buffers[], void *cl_arg)
  342. { }
  343. struct starpu_codelet cl_switch =
  344. {
  345. .cpu_funcs = {empty},
  346. .nbuffers = STARPU_VARIABLE_NBUFFERS,
  347. };
  348. ret = starpu_task_insert(&cl_switch, STARPU_RW, handle,
  349. STARPU_W, vert_handle[0],
  350. STARPU_W, vert_handle[1],
  351. 0);
  352. \endcode
  353. The execution of the <c>switch</c> task will get back the matrix data into the
  354. main memory, and thus the vertical slices will get the updated value there.
  355. Again, we prefer to make sure that we don't accidentally access the matrix through the whole-matrix handle:
  356. \code{.c}
  357. starpu_data_invalidate_submit(handle);
  358. \endcode
  359. And now we can start using vertical slices, etc.
  360. \section DefiningANewDataFilter Defining A New Data Filter
  361. StarPU provides a series of predefined filters in \ref API_Data_Partition, but
  362. additional filters can be defined by the application. The principle is that the
  363. filter function just fills the memory location of the <c>i-th</c> subpart of a data.
  364. Examples are provided in <c>src/datawizard/interfaces/*_filters.c</c>,
  365. and see \ref starpu_data_filter::filter_func for the details.
  366. \section DataReduction Data Reduction
  367. In various cases, some piece of data is used to accumulate intermediate
  368. results. For instances, the dot product of a vector, maximum/minimum finding,
  369. the histogram of a photograph, etc. When these results are produced along the
  370. whole machine, it would not be efficient to accumulate them in only one place,
  371. incurring data transmission each and access concurrency.
  372. StarPU provides a mode ::STARPU_REDUX, which permits to optimize
  373. this case: it will allocate a buffer on each memory node, and accumulate
  374. intermediate results there. When the data is eventually accessed in the normal
  375. mode ::STARPU_R, StarPU will collect the intermediate results in just one
  376. buffer.
  377. For this to work, the user has to use the function
  378. starpu_data_set_reduction_methods() to declare how to initialize these
  379. buffers, and how to assemble partial results.
  380. For instance, <c>cg</c> uses that to optimize its dot product: it first defines
  381. the codelets for initialization and reduction:
  382. \code{.c}
  383. struct starpu_codelet bzero_variable_cl =
  384. {
  385. .cpu_funcs = { bzero_variable_cpu },
  386. .cpu_funcs_name = { "bzero_variable_cpu" },
  387. .cuda_funcs = { bzero_variable_cuda },
  388. .nbuffers = 1,
  389. }
  390. static void accumulate_variable_cpu(void *descr[], void *cl_arg)
  391. {
  392. double *v_dst = (double *)STARPU_VARIABLE_GET_PTR(descr[0]);
  393. double *v_src = (double *)STARPU_VARIABLE_GET_PTR(descr[1]);
  394. *v_dst = *v_dst + *v_src;
  395. }
  396. static void accumulate_variable_cuda(void *descr[], void *cl_arg)
  397. {
  398. double *v_dst = (double *)STARPU_VARIABLE_GET_PTR(descr[0]);
  399. double *v_src = (double *)STARPU_VARIABLE_GET_PTR(descr[1]);
  400. cublasaxpy(1, (double)1.0, v_src, 1, v_dst, 1);
  401. cudaStreamSynchronize(starpu_cuda_get_local_stream());
  402. }
  403. struct starpu_codelet accumulate_variable_cl =
  404. {
  405. .cpu_funcs = { accumulate_variable_cpu },
  406. .cpu_funcs_name = { "accumulate_variable_cpu" },
  407. .cuda_funcs = { accumulate_variable_cuda },
  408. .nbuffers = 1,
  409. }
  410. \endcode
  411. and attaches them as reduction methods for its handle <c>dtq</c>:
  412. \code{.c}
  413. starpu_variable_data_register(&dtq_handle, -1, NULL, sizeof(type));
  414. starpu_data_set_reduction_methods(dtq_handle, &accumulate_variable_cl, &bzero_variable_cl);
  415. \endcode
  416. and <c>dtq_handle</c> can now be used in mode ::STARPU_REDUX for the
  417. dot products with partitioned vectors:
  418. \code{.c}
  419. for (b = 0; b < nblocks; b++)
  420. starpu_task_insert(&dot_kernel_cl,
  421. STARPU_REDUX, dtq_handle,
  422. STARPU_R, starpu_data_get_sub_data(v1, 1, b),
  423. STARPU_R, starpu_data_get_sub_data(v2, 1, b),
  424. 0);
  425. \endcode
  426. During registration, we have here provided <c>NULL</c>, i.e. there is
  427. no initial value to be taken into account during reduction. StarPU
  428. will thus only take into account the contributions from the tasks
  429. <c>dot_kernel_cl</c>. Also, it will not allocate any memory for
  430. <c>dtq_handle</c> before tasks <c>dot_kernel_cl</c> are ready to run.
  431. If another dot product has to be performed, one could unregister
  432. <c>dtq_handle</c>, and re-register it. But one can also call
  433. starpu_data_invalidate_submit() with the parameter <c>dtq_handle</c>,
  434. which will clear all data from the handle, thus resetting it back to
  435. the initial status <c>register(NULL)</c>.
  436. The example <c>cg</c> also uses reduction for the blocked gemv kernel,
  437. leading to yet more relaxed dependencies and more parallelism.
  438. ::STARPU_REDUX can also be passed to starpu_mpi_task_insert() in the MPI
  439. case. This will however not produce any MPI communication, but just pass
  440. ::STARPU_REDUX to the underlying starpu_task_insert(). It is up to the
  441. application to call starpu_mpi_redux_data(), which posts tasks which will
  442. reduce the partial results among MPI nodes into the MPI node which owns the
  443. data. For instance, some hypothetical application which collects partial results
  444. into data <c>res</c>, then uses it for other computation, before looping again
  445. with a new reduction:
  446. \code{.c}
  447. for (i = 0; i < 100; i++)
  448. {
  449. starpu_mpi_task_insert(MPI_COMM_WORLD, &init_res, STARPU_W, res, 0);
  450. starpu_mpi_task_insert(MPI_COMM_WORLD, &work, STARPU_RW, A, STARPU_R, B, STARPU_REDUX, res, 0);
  451. starpu_mpi_redux_data(MPI_COMM_WORLD, res);
  452. starpu_mpi_task_insert(MPI_COMM_WORLD, &work2, STARPU_RW, B, STARPU_R, res, 0);
  453. }
  454. \endcode
  455. \section DataCommute Commute Data Access
  456. By default, the implicit dependencies computed from data access use the
  457. sequential semantic. Notably, write accesses are always serialized in the order
  458. of submission. In some applicative cases, the write contributions can actually
  459. be performed in any order without affecting the eventual result. In this case
  460. it is useful to drop the strictly sequential semantic, to improve parallelism
  461. by allowing StarPU to reorder the write accesses. This can be done by using
  462. the ::STARPU_COMMUTE data access flag. Accesses without this flag will however
  463. properly be serialized against accesses with this flag. For instance:
  464. \code{.c}
  465. starpu_task_insert(&cl1, STARPU_R, h, STARPU_RW, handle, 0);
  466. starpu_task_insert(&cl2, STARPU_R, handle1, STARPU_RW|STARPU_COMMUTE, handle, 0);
  467. starpu_task_insert(&cl2, STARPU_R, handle2, STARPU_RW|STARPU_COMMUTE, handle, 0);
  468. starpu_task_insert(&cl3, STARPU_R, g, STARPU_RW, handle, 0);
  469. \endcode
  470. The two tasks running <c>cl2</c> will be able to commute: depending on whether the
  471. value of <c>handle1</c> or <c>handle2</c> becomes available first, the corresponding task
  472. running <c>cl2</c> will start first. The task running <c>cl1</c> will however always be run
  473. before them, and the task running <c>cl3</c> will always be run after them.
  474. If a lot of tasks use the commute access on the same set of data and a lot of
  475. them are ready at the same time, it may become interesting to use an arbiter,
  476. see \ref ConcurrentDataAccess.
  477. \section ConcurrentDataAccess Concurrent Data Accesses
  478. When several tasks are ready and will work on several data, StarPU is faced with
  479. the classical Dining Philosophers problem, and has to determine the order in
  480. which it will run the tasks.
  481. Data accesses usually use sequential ordering, so data accesses are usually
  482. already serialized, and thus by default StarPU uses the Dijkstra solution which
  483. scales very well in terms of overhead: tasks will just acquire data one by one
  484. by data handle pointer value order.
  485. When sequential ordering is disabled or the ::STARPU_COMMUTE flag is used, there
  486. may be a lot of concurrent accesses to the same data, and the Dijkstra solution
  487. gets only poor parallelism, typically in some pathological cases which do happen
  488. in various applications. In this case, one can use a data access arbiter, which
  489. implements the classical centralized solution for the Dining Philosophers
  490. problem. This is more expensive in terms of overhead since it is centralized,
  491. but it opportunistically gets a lot of parallelism. The centralization can also
  492. be avoided by using several arbiters, thus separating sets of data for which
  493. arbitration will be done. If a task accesses data from different arbiters, it
  494. will acquire them arbiter by arbiter, in arbiter pointer value order.
  495. See the <c>tests/datawizard/test_arbiter.cpp</c> example.
  496. Arbiters however do not support the ::STARPU_REDUX flag yet.
  497. \section TemporaryBuffers Temporary Buffers
  498. There are two kinds of temporary buffers: temporary data which just pass results
  499. from a task to another, and scratch data which are needed only internally by
  500. tasks.
  501. \subsection TemporaryData Temporary Data
  502. Data can sometimes be entirely produced by a task, and entirely consumed by
  503. another task, without the need for other parts of the application to access
  504. it. In such case, registration can be done without prior allocation, by using
  505. the special memory node number <c>-1</c>, and passing a zero pointer. StarPU will
  506. actually allocate memory only when the task creating the content gets scheduled,
  507. and destroy it on unregistration.
  508. In addition to this, it can be tedious for the application to have to unregister
  509. the data, since it will not use its content anyway. The unregistration can be
  510. done lazily by using the function starpu_data_unregister_submit(),
  511. which will record that no more tasks accessing the handle will be submitted, so
  512. that it can be freed as soon as the last task accessing it is over.
  513. The following code examplifies both points: it registers the temporary
  514. data, submits three tasks accessing it, and records the data for automatic
  515. unregistration.
  516. \code{.c}
  517. starpu_vector_data_register(&handle, -1, 0, n, sizeof(float));
  518. starpu_task_insert(&produce_data, STARPU_W, handle, 0);
  519. starpu_task_insert(&compute_data, STARPU_RW, handle, 0);
  520. starpu_task_insert(&summarize_data, STARPU_R, handle, STARPU_W, result_handle, 0);
  521. starpu_data_unregister_submit(handle);
  522. \endcode
  523. The application may also want to see the temporary data initialized
  524. on the fly before being used by the task. This can be done by using
  525. starpu_data_set_reduction_methods() to set an initialization codelet (no redux
  526. codelet is needed).
  527. \subsection ScratchData Scratch Data
  528. Some kernels sometimes need temporary data to achieve the computations, i.e. a
  529. workspace. The application could allocate it at the start of the codelet
  530. function, and free it at the end, but this would be costly. It could also
  531. allocate one buffer per worker (similarly to \ref HowToInitializeAComputationLibraryOnceForEachWorker),
  532. but this would
  533. make them systematic and permanent. A more optimized way is to use
  534. the data access mode ::STARPU_SCRATCH, as examplified below, which
  535. provides per-worker buffers without content consistency. The buffer is
  536. registered only once, using memory node <c>-1</c>, i.e. the application didn't allocate
  537. memory for it, and StarPU will allocate it on demand at task execution.
  538. \code{.c}
  539. starpu_vector_data_register(&workspace, -1, 0, sizeof(float));
  540. for (i = 0; i < N; i++)
  541. starpu_task_insert(&compute, STARPU_R, input[i], STARPU_SCRATCH, workspace, STARPU_W, output[i], 0);
  542. \endcode
  543. StarPU will make sure that the buffer is allocated before executing the task,
  544. and make this allocation per-worker: for CPU workers, notably, each worker has
  545. its own buffer. This means that each task submitted above will actually have its
  546. own workspace, which will actually be the same for all tasks running one after
  547. the other on the same worker. Also, if for instance memory becomes scarce,
  548. StarPU will notice that it can free such buffers easily, since the content does
  549. not matter.
  550. The example <c>examples/pi</c> uses scratches for some temporary buffer.
  551. \section TheMultiformatInterface The Multiformat Interface
  552. It may be interesting to represent the same piece of data using two different
  553. data structures: one only used on CPUs, and one only used on GPUs.
  554. This can be done by using the multiformat interface. StarPU
  555. will be able to convert data from one data structure to the other when needed.
  556. Note that the scheduler <c>dmda</c> is the only one optimized for this
  557. interface. The user must provide StarPU with conversion codelets:
  558. \snippet multiformat.c To be included. You should update doxygen if you see this text.
  559. Kernels can be written almost as for any other interface. Note that
  560. ::STARPU_MULTIFORMAT_GET_CPU_PTR shall only be used for CPU kernels. CUDA kernels
  561. must use ::STARPU_MULTIFORMAT_GET_CUDA_PTR, and OpenCL kernels must use
  562. ::STARPU_MULTIFORMAT_GET_OPENCL_PTR. ::STARPU_MULTIFORMAT_GET_NX may
  563. be used in any kind of kernel.
  564. \code{.c}
  565. static void
  566. multiformat_scal_cpu_func(void *buffers[], void *args)
  567. {
  568. struct point *aos;
  569. unsigned int n;
  570. aos = STARPU_MULTIFORMAT_GET_CPU_PTR(buffers[0]);
  571. n = STARPU_MULTIFORMAT_GET_NX(buffers[0]);
  572. ...
  573. }
  574. extern "C" void multiformat_scal_cuda_func(void *buffers[], void *_args)
  575. {
  576. unsigned int n;
  577. struct struct_of_arrays *soa;
  578. soa = (struct struct_of_arrays *) STARPU_MULTIFORMAT_GET_CUDA_PTR(buffers[0]);
  579. n = STARPU_MULTIFORMAT_GET_NX(buffers[0]);
  580. ...
  581. }
  582. \endcode
  583. A full example may be found in <c>examples/basic_examples/multiformat.c</c>.
  584. \section DefiningANewDataInterface Defining A New Data Interface
  585. Let's define a new data interface to manage complex numbers.
  586. \code{.c}
  587. /* interface for complex numbers */
  588. struct starpu_complex_interface
  589. {
  590. double *real;
  591. double *imaginary;
  592. int nx;
  593. };
  594. \endcode
  595. Registering such a data to StarPU is easily done using the function
  596. starpu_data_register(). The last
  597. parameter of the function, <c>interface_complex_ops</c>, will be
  598. described below.
  599. \code{.c}
  600. void starpu_complex_data_register(starpu_data_handle_t *handle,
  601. unsigned home_node, double *real, double *imaginary, int nx)
  602. {
  603. struct starpu_complex_interface complex =
  604. {
  605. .real = real,
  606. .imaginary = imaginary,
  607. .nx = nx
  608. };
  609. if (interface_complex_ops.interfaceid == STARPU_UNKNOWN_INTERFACE_ID)
  610. {
  611. interface_complex_ops.interfaceid = starpu_data_interface_get_next_id();
  612. }
  613. starpu_data_register(handleptr, home_node, &complex, &interface_complex_ops);
  614. }
  615. \endcode
  616. Different operations need to be defined for a data interface through
  617. the type starpu_data_interface_ops. We only define here the basic
  618. operations needed to run simple applications. The source code for the
  619. different functions can be found in the file
  620. <c>examples/interface/complex_interface.c</c>, the details of the hooks to be
  621. provided are documented in \ref starpu_data_interface_ops .
  622. \code{.c}
  623. static struct starpu_data_interface_ops interface_complex_ops =
  624. {
  625. .register_data_handle = complex_register_data_handle,
  626. .allocate_data_on_node = complex_allocate_data_on_node,
  627. .copy_methods = &complex_copy_methods,
  628. .get_size = complex_get_size,
  629. .footprint = complex_footprint,
  630. .interfaceid = STARPU_UNKNOWN_INTERFACE_ID,
  631. .interface_size = sizeof(struct starpu_complex_interface),
  632. };
  633. \endcode
  634. Functions need to be defined to access the different fields of the
  635. complex interface from a StarPU data handle.
  636. \code{.c}
  637. double *starpu_complex_get_real(starpu_data_handle_t handle)
  638. {
  639. struct starpu_complex_interface *complex_interface =
  640. (struct starpu_complex_interface *) starpu_data_get_interface_on_node(handle, STARPU_MAIN_RAM);
  641. return complex_interface->real;
  642. }
  643. double *starpu_complex_get_imaginary(starpu_data_handle_t handle);
  644. int starpu_complex_get_nx(starpu_data_handle_t handle);
  645. \endcode
  646. Similar functions need to be defined to access the different fields of the
  647. complex interface from a <c>void *</c> pointer to be used within codelet
  648. implemetations.
  649. \snippet complex.c To be included. You should update doxygen if you see this text.
  650. Complex data interfaces can then be registered to StarPU.
  651. \code{.c}
  652. double real = 45.0;
  653. double imaginary = 12.0;
  654. starpu_complex_data_register(&handle1, STARPU_MAIN_RAM, &real, &imaginary, 1);
  655. starpu_task_insert(&cl_display, STARPU_R, handle1, 0);
  656. \endcode
  657. and used by codelets.
  658. \code{.c}
  659. void display_complex_codelet(void *descr[], void *_args)
  660. {
  661. int nx = STARPU_COMPLEX_GET_NX(descr[0]);
  662. double *real = STARPU_COMPLEX_GET_REAL(descr[0]);
  663. double *imaginary = STARPU_COMPLEX_GET_IMAGINARY(descr[0]);
  664. int i;
  665. for(i=0 ; i<nx ; i++)
  666. {
  667. fprintf(stderr, "Complex[%d] = %3.2f + %3.2f i\n", i, real[i], imaginary[i]);
  668. }
  669. }
  670. \endcode
  671. The whole code for this complex data interface is available in the
  672. directory <c>examples/interface/</c>.
  673. \section SpecifyingATargetNode Specifying A Target Node For Task Data
  674. When executing a task on a GPU for instance, StarPU would normally copy all the
  675. needed data for the tasks on the embedded memory of the GPU. It may however
  676. happen that the task kernel would rather have some of the datas kept in the
  677. main memory instead of copied in the GPU, a pivoting vector for instance.
  678. This can be achieved by setting the starpu_codelet::specific_nodes flag to
  679. <c>1</c>, and then fill the starpu_codelet::nodes array (or starpu_codelet::dyn_nodes when
  680. starpu_codelet::nbuffers is greater than \ref STARPU_NMAXBUFS) with the node numbers
  681. where data should be copied to, or ::STARPU_SPECIFIC_NODE_LOCAL to let
  682. StarPU copy it to the memory node where the task will be executed.
  683. ::STARPU_SPECIFIC_NODE_CPU can also be used to request data to be
  684. put in CPU-accessible memory (and let StarPU choose the NUMA node).
  685. ::STARPU_SPECIFIC_NODE_FAST and ::STARPU_SPECIFIC_NODE_SLOW can also be
  686. used
  687. For instance,
  688. with the following codelet:
  689. \code{.c}
  690. struct starpu_codelet cl =
  691. {
  692. .cuda_funcs = { kernel },
  693. .nbuffers = 2,
  694. .modes = {STARPU_RW, STARPU_RW},
  695. .specific_nodes = 1,
  696. .nodes = {STARPU_SPECIFIC_NODE_CPU, STARPU_SPECIFIC_NODE_LOCAL},
  697. };
  698. \endcode
  699. the first data of the task will be kept in the CPU memory, while the second
  700. data will be copied to the CUDA GPU as usual. A working example is available in
  701. <c>tests/datawizard/specific_node.c</c>
  702. With the following codelet:
  703. \code{.c}
  704. struct starpu_codelet cl =
  705. {
  706. .cuda_funcs = { kernel },
  707. .nbuffers = 2,
  708. .modes = {STARPU_RW, STARPU_RW},
  709. .specific_nodes = 1,
  710. .nodes = {STARPU_SPECIFIC_NODE_LOCAL, STARPU_SPECIFIC_NODE_SLOW},
  711. };
  712. \endcode
  713. The first data will be copied into fast (but probably size-limited) local memory
  714. while the second data will be left in slow (but large) memory. This makes sense
  715. when the kernel does not make so many accesses to the second data, and thus data
  716. being remote e.g. over a PCI bus is not a performance problem, and avoids
  717. filling the fast local memory with data which does not need the performance.
  718. */