Databases
StackOverflow2013
Postsdbo
PO
 

Note that there are some explanatory texts on larger screens.

plurals
  1. PO
    primarykey
    data
    text
    <p>There are some cases where you can get programs to run at the full potential speed on the GPU with very little porting work from a plain CPU version, and this might be one of them.</p> <p>If it's possible for you to have a function like this:</p> <pre><code>void process_single_video_frame(void* part_of_frame) { // initialize variables ... intermediate_result_1 = function1(part_of_frame); intermediate_result_2 = function2(intermediate_result_1); intermediate_result_3 = function3(intermediate_result_2); store_results(intermediate_result_3); } </code></pre> <p><strong>and</strong> you can process many part_of_frames at the same time. Say, a few thousand,</p> <p><strong>and</strong> <code>function1()</code>, <code>function2()</code> and <code>function3()</code> go through pretty much the same code paths (that is, the program flow does not depend heavily on the contents of the frame),</p> <p>then, local memory may do all the work for you. Local memory is a type of memory that is stored in global memory. It is different from global memory in a subtle, yet profound way... The memory is simply interleaved in such a way that adjacent threads will access adjacent 32 bit words, enabling the memory access to be fully coalesced if the threads all read from the same location of their local memory arrays.</p> <p>The flow of your program would be that you start out by copying <code>part_of_frame</code> to a local array and prepare other local arrays for intermediate results. You then pass pointers to the local arrays between the various functions in your code.</p> <p>Some pseudocode:</p> <pre><code>const int size_of_one_frame_part = 1000; __global__ void my_kernel(int* all_parts_of_frames) { int i = blockIdx.x * blockDim.x + threadIdx.x; int my_local_array[size_of_one_frame_part]; memcpy(my_local_array, all_parts_of_frames + i * size_of_one_frame_part); int local_intermediate_1[100]; function1(local_intermediate_1, my_local_array); ... } __device__ void function1(int* dst, int* src) { ... } </code></pre> <p>In summary, this approach may let you use your CPU functions pretty much unchanged, as the parallelism does not come from creating parallelized versions of your functions, but instead by running the entire chain of functions in parallel. And this again is made possible by the hardware support for interleaving the memory in local arrays.</p> <p>Notes:</p> <ul> <li><p>The initial copy of the <code>part_of_frame</code> from global to local memory is not coalesced, but hopefully, you will have enough calculations to hide that.</p></li> <li><p>On devices of compute capability &lt;= 1.3, there is only 16KiB of local memory available per thread, which may not be enough for your <code>part_of_frame</code> and the other intermediate data. But on compute capability >= 2.0, this has bee expanded to 512KiB, which should be plenty.</p></li> </ul>
    singulars
    1. This table or related slice is empty.
    1. POProper way to write kernel functions in CUDA?
      singulars
      1. PTQuestion
    1. PTAnswer
    1. USRoger Dahl
    1. USRoger Dahl
    plurals
    1. This table or related slice is empty.
    1. This table or related slice is empty.
    1. POProper way to write kernel functions in CUDA?
      singulars
      1. PTQuestion
    1. This table or related slice is empty.
    1. VO
      singulars
      1. PO
      1. This table or related slice is empty.
      1. VTUpMod
    2. VO
      singulars
      1. PO
      1. This table or related slice is empty.
      1. VTUpMod
    3. VO
      singulars
      1. PO
      1. This table or related slice is empty.
      1. VTAcceptedByOriginator
    1. COThis is a very interesting approach, however if I understand your your method correctly I do not think I can do this. I first start with a 512x512 raw video file. From that, there is a 128x128 section I analyze and do all the functions on. From that 128 by 128 section, certain tasks are dependent on the entire image, so I would not be able to independently take a 16x16 square and do manipulations on that, and then just add that at the end.
      singulars
      1. PO
      1. USShawn Tabrizi
    2. COOkay after rereading your idea, I think it might work, but I want to clarify to make sure I am not taking your idea out of proportion. My C program currently runs on 1 frame, using a single CPU thread (as far as I can tell). Would I be able to then pass #MAXBLOCKS*#MAXTHREADS number of images to my program, and have each thread go though my C program once on the frame? Which means that it will be slow to get a result, but I would get like 60,000 frames processed in one go? as in, use int i = blockIdx.x * blockDim.x + threadIdx.x; to determine which image to processes, and then crunch it.
      singulars
      1. PO
      1. USShawn Tabrizi
    3. COAlso should I be worried about the sheer magnitude of runtime variables and memory allocation at this scale? Assuming I process ~60,000 images, each on one thread, this alone is around 1gb of data (128x128 8 bit image). Then I will be also making 60,000 copies of all the run time variables, and allocating them all in the kernel, which i have read is not a good idea.
      singulars
      1. PO
      1. USShawn Tabrizi
 

Querying!

 
Guidance
A row detail

Detail views are divided into sections. All the information in the data section comes from columns in the selected row. The other sections display data from other, related rows.

Related data can be related in a to-one or a to-many fashion. Captions of data related in a to-many fashion link to a list view showing a filtered view of the table.

Try moving around until you find a non-empty to-many entry and click on the label to get to one. You can move back to the root by clicking on the database name in the header.

SQuiL has stopped working due to an internal error.

If you are curious you may find further information in the browser console, which is accessible through the devtools (F12).

Reload