cuda - atomicAdd() for double on GPU -


i doing project on gpu, , have use atomicadd() double, because cuda not support double, use code below, nvidia provide.

__device__ double atomicadd(double* address, double val) {     unsigned long long int* address_as_ull =                                           (unsigned long long int*)address;     unsigned long long int old = *address_as_ull, assumed;     {         assumed = old;         old = atomiccas(address_as_ull, assumed,                          __double_as_longlong(val +                          __longlong_as_double(assumed)));     } while (assumed != old);     return __longlong_as_double(old); }

now want know why implement require loop, while (assumed!=old)

basically because implementation requires load, can't performed atomically. compare-and-swap operation atomic version of

(*address == assumed) ? (assumed + val) : *address

there no guarantee the value @ address won't change between cycle value loaded address , cycle atomiccas call used store updated value. if happens, value @ address won't updated. therefore loop ensures 2 operations repeated until there no change of value @ address between read , compare-and-swap operation, implies update took place.


Comments

Post a Comment

Popular posts from this blog

Why does Ruby on Rails generate add a blank line to the end of a file? -

running rails generate controller foo home results in: class foocontroller < application controller def home end end (there's nothing on line in actual file, blank line) is there purpose blank line? it's common use newline indicate end of line, last line. many editors (like vi, use) add in newline silently. consensus text files (especially in unix world) should end in newline, , there have been problems historically if wasn't present. why should text files end newline? the tool used count lines in file "wc" counts newlines in file, without trailing newline show 3 instead of 4. it improves readability of templates used in generator. consider: https://github.com/rails/rails/blob/master/railties/lib/rails/generators/rails/controller/templates/controller.rb to remove trailing newline, template have last line of: end<% end -%> instead of: end <% end -%> that seems less readable me. the discussion below refers ...
Read more

keyboard - Smiles and long press feature in Android -

i've created android keyboard. how add features inserting smilies, , using long keypresses display symbols? to create long-press behavior allows select several options pop-up menu, follow guide (specifically pop-up menu section): http://developer.android.com/guide/topics/ui/menus.html smileys (also known emoticons) combination of keys, understood programs , parsed same way. have do, when clicks smiley face on keyboard, simulate pressing of key combination matching emoticon. example, when clicks smiley, should simulate key presses ':' , ')' :). here's list of emoticons: http://en.wikipedia.org/wiki/list_of_emoticons
Read more

node.js - Bad Request - node js ajax post -

i'm using express 3.0 , node v0.11.0. have button submits form , i'm using ajax post data json object node server. client code is: $('#contact').submit(function(e) { console.log('submit called'); var formdata = json.stringify($('form').serializeobject()); console.log(formdata); $.ajax({ url: "http://localhost:3000/savedata/", type: "post", datatype: 'json', data: {objectdata: formdata}, contenttype: "application/json", complete: function() { console.log('process complete'); }, success: function(data) { console.log('process sucess'); }, error: function() { console.log('process error'); }, }); return false; }); then on server: var express = require('express'); var app = express(); app.set('views', __dirname + '/views'); app.set('view engine', 'jade'); app.use(express.bodyparser());...
Read more