Hello,
I’m currently assessing the possibility of using weird CUDA deadlocking behaviors in order to generate random numbers…
I’m using this kernel in order to generate deadlocks, on my GTX260… with a block size of 512 threads:
[codebox]
extern shared unsigned int sh_arr_a;
global void
kernel_dl( void* data,
unsigned int dataSize)
{
const unsigned int tid = threadIdx.x;
const unsigned int bid = blockIdx.x;
const unsigned int bdim = blockDim.x;
const unsigned int ind = bid * bdim + tid;
if( ind > dataSize-1 )
return;
unsigned int * arr = (unsigned int *)data;
arr[ind]=0;
(sh_arr_a)[ tid ] = 0;
for( int i = 0; i < 1; ++i )
{
sh_arr_a[ tid ] = (sh_arr_a[ (tid+96)%512 ]+
sh_arr_a[ (tid+64)%512 ]+
sh_arr_a[ (tid+32)%512 ]);
//__syncthreads();
}
__syncthreads();
arr[ind] = sh_arr_a[ tid ];
}[/codebox]
The output is :
[codebox]0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
1056964608
1056964608
1056964608
1056964608
1056964608
1056964608
1056964608
1056964608
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
1065353216
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
999985384
1007740890
1141538816
1135050752
1111097344
1111097344
1111097344
1111097344
1022330940
1022330940
1025748521
1025748521
1025748521
1025748521
1025748521
1025748521
1053923125
1053923125
1131347968
1131347968
1132138424
1132138424
1132806144
1132806144
0
1154613248
0
1154613248
1154613248
0
0
1154613248
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
0
[/codebox]
As you can see, there are some weird values appearing here and there…
After 512 iterations… We obtain this:
[codebox]66011136
4196130816
2970542080
3135389696
2809077760
1940889600
73646080
941834240
3474243904
613957952
1658011968
223330624
1364082688
2798764032
3894689792
2460008448
3454509056
735035392
1111572480
3831046144
4276433260
1988473196
2365010284
358003052
478230784
3905001728
210868480
1079064832
1655570432
1894776832
3044540416
2805334016
3461005312
129785856
634413056
3965632512
1516290048
285556736
2475016192
3705749504
3747464896
3922446016
3602040512
3427059392
2219704320
2394685440
2326298624
2151317504
1398808576
20496384
1943355392
3321667584
1810238538
591957066
2514816074
3733097546
3725992832
3098100608
4089004928
421929856
1602879488
3937402880
2505572352
171048960
3818438656
145399808
3029311488
2407383040
4236763136
1931567104
3657252864
1667481600
4244882624
1923105984
2639040
2324415680
1669775360
3642966016
2141405184
168214528
1782218752
1948319744
2948988928
2782887936
1527507955
2191911923
3192581107
2528177139
2175619136
753307712
1395036224
2817347648
618921984
2017460224
333447168
3229876224
2075197440
2139455488
2145353728
2081095680
1191788544
1204502528
3116433408
3103719424
2528694720
985256384
3696677312
945148352
4088692736
2545254400
1884487680
3427926016
2716860416
614727680
2402942976
210108416
2922903097
3104306745
597554745
416151097
1773992128
354056384
119175360
1539111104
884473856
3439329280
2315255808
4055367680
1862336512
2139897856
4049682432
3772121088
382705664
1444454400
4291264512
3229515776
1268854848
1014083648
4197691456
157495360
756203520
501432320
2595373056
2850144256
1194696704
1752416256
886390784
328671232
1654794782
3885672990
3019647518
788769310
4071348864
1990736512
2574793344
360438400
99483648
2775580672
4151836672
1475739648
2779234304
2377080832
1693409280
2095562752
2073051136
786743296
2701115392
3987423232
3042325568
2796172352
338048064
584201280
1075445760
829292544
996933632
1243086848
2929819648
51183616
2290679808
874348544
249182745
1619540505
3859036697
2488678937
2624069824
3809943744
1104617664
4213711040
2742026240
2344615936
1822425088
2219835392
2196013056
1486225408
2774335488
3484123136
2992988160
612065280
2228158464
314114048
2536986048
3882669504
282842560
3232126400
1774010368
3119693824
2770288640
1424605184
3575980032
4183801856
875151360
267329536
101732004
2533019300
3519336100
1088048804
1308668672
2886939392
1411592960
4128289536
2826436608
2368733184
4166516736
329252864
1182056448
2571558912
3876904960
2487402496
3906600960
45645824
3656957952
3222945792
4279477824
2925176384
3576800832
636134976
2201714688
847413248
3123118080
182452224
3278831616
1812398080
2015526912
3481960448
1303880033
4028080481
4231209313
1507008865
138610368
3883787968
58582720
608372416
3120955392
2357198848
2457862144
3221618688
3696132096
2614263808
3632103424
419004416
3763396608
129032192
1110966272
450363392
520828992
283293760
773371968
1010907200
1394688000
1157152768
3693461504
3930996736
369975296
2644918272
3694968832
1420025856
3138538004
3648375316
403458580
4188588564
1176790784
1334183680
3929409280
3772016384
1089601536
1913651200
3787980800
2963931136
1971650560
3406807040
3963535360
2528378880
743890944
1717952512
2363785216
1389723648
3748266368
3478356352
2618655104
2888565120
4195368960
3925458944
467189760
737099776
539238400
625459200
3740975104
3654754304
245717888
590601088
3706116992
3361233792
1985060864
2757648384
4008599552
3236012032
2666266624
3417309184
3843555328
3092512768
3414441984
1795219456
1768022016
3387244544
2094522368
1368940544
4018913280
449527808
2028177344
2448558016
1875183552
1454802880
1928249344
2348630016
1705230336
1284849664
2907791360
2769174528
751616000
890232832
2720039511
2165572183
148013655
702480983
1676301120
4063097664
1457386304
3365557056
1938948096
1611661312
2831679488
3158966272
1965457408
1826816000
2164260864
2302902272
1287528448
2636652544
3356426240
2007302144
1137832320
1424126336
119697792
4128371072
2792669184
3078963200
2773827584
2487533568
2276884480
1915551744
1462239232
1823571968
964813707
3814450059
3361137547
511501195
3899773504
2094043712
3406860864
917623360
2946236416
2495610880
2489319424
2939944960
2614181888
444784640
1927405568
4096802816
48734208
3109822464
1472577536
2706456576
130945472
2307625408
635474368
2753761728
1995128832
4171808768
3569762304
1393082368
3002040320
454950912
608796672
3155886080
2355738424
757315384
911161144
2509584184
3699046912
1691474432
521263616
2528836096
2735079424
4173332480
3328180224
1889927168
112066560
569966592
369819648
4206886912
3155640320
2182823936
2685665280
3658481664
382582976
1919074496
2899689664
1363198144
3993796608
1235320832
2454454272
917962752
3452895232
667811840
2154299392
644415488
3789182309
1238783333
2725270885
980702565
1029636544
348488128
373653952
1054802368
1896349696
2139095040
2873098240
2630352896
3725557760
2124562432
3421782016
727810048
2955657216
1904918528
512000000
1562738688
3820273280
713408128
1061797504
4168662656
3751641088
644775936
4205084672
3016982528
285548544
3657547776
1190576128
2113544192
208679111
811774151
2639769799
2036674759
911624000
4053501760
1990952768
3144042304
2652110848
967835648
2102919168
3787194368
198541312
2093613056
4111794176
2216722432
2925256704
3973439488
2258305024
1210122240
893531776
471971456
96712320
518272640
167854080
4041261056
2279063552
2700623872
4022165504
2902016000
2430615552
3550765056
2711941905
2526311185
2054910737
2240541457
1002367680
1345366720
1986046656
1643047616
3688628224
3725590528
847773696
810811392
[/codebox]
And if we normalize these numbers to [0,1], and we plot them on a graph, we obtain this:
So I was actually thinking I could use this method as a basis for a PRNG…
My question is: Is this behavior normal ? Will it present on every CUDA-capable cards ? What’s actually happening ?
Thank you.