[Campaign-news] New k-centers

[Marc Sosnick]

Kai:

Aaah...I see, sorry misunderstood what you were getting at.

As far as I know there's no way to call the kernel directly from the C  
code.  The  __global__ declaration specifier is a CUDA C extension  
provided through nvcc, so there'd be no way to directly call kernel  
code from  a c program except going through a nvcc preprocess anyway.

To directly control a kernel we could change our overall approach and  
use the CUDA driver API.  This would *significantly* complicate the  
code, so we'd have to think about the tradeoffs.  We can talk about  
this tonight if you like.

See you later!

Marc





On Feb 11, 2010, at 5:46 PM, Kai Kohlhoff wrote:

> Hey Marc,
>
> Yes, pretty cool, and the way it it supposed to work with regular C+ 
> + (I know that).
>
> Could you please try this with, e.g.:
> __global__ void assignCluster(int N, int g, int *assignments_d,  
> float *distances_d, float *dist_d);
>
> That's where the standard fails.
>
> I would like to be able to call the kernels directly from any  
> program that uses them, rather than having to rely on a 'middle man'  
> such as kcentersGPU().  It increases flexibility.
>
> Cheers,
> Kai
>
>
>
>
> On Feb 11, 2010, at 5:37 PM, Marc Sosnick wrote:
>
>> Kai:
>>
>>>> For your example, the kcenterGPU.cu file should not be included,  
>>>> but rather linked in during the link phase, and instead a  
>>>> kcenterGPU.h file included in the users' projects.
>>>
>>> Hi Marc,
>>>
>>> That all sounds good and I am looking forward to talking more  
>>> about this tomorrow.
>>>
>>> I just have a quick remark to the last bit of your email.  I am  
>>> not sure that using an extra .h file is even possible.  Do you  
>>> have an example? For all I know, it is not supported that kernels  
>>> be split into a header and a source code file (e.g. keywords like  
>>> __global__ are not supported in a separate header declaration).   
>>> The .cu in principle IS the .h file containing both declaration  
>>> and definition.  It does work for your timing object, since you  
>>> are not actually implementing GPU kernels, but I would be (very  
>>> positively!) surprised if this would work for kcentersGPU.  If you  
>>> have a solution I am very eager to hear about it.
>>>
>>
>> It actually works like a charm.  All the .h file does is have  
>> forward declarations of functions that allows the compiler to  
>> continue compiling.  The compiler essentially says "here's where  
>> this function call goes, let's wait to the linker to put the  
>> address in."  NVCC precompiles the .cu file, then gcc links the .o  
>> file in which contains the actual definition of the function.  So,  
>> all the .h file contains is essentially the line that you're using  
>> now  (for example):
>>
>>       void kcentersGPU(int N, int K, float *data, int *assignments,  
>> float *distances, float *generators, int seed);
>>
>> Pretty cool, eh?
>>
>> Marc
>> _______________________________________________
>> Campaign-news mailing list
>> Campaign-news at simtk.org
>> https://simtk.org/mailman/listinfo/campaign-news
>
> -----------------------------------------------------
> Kai Kohlhoff, PhD
> Stanford University
> School of Medicine, Bioengineering
> Stanford, CA 94305-5448, USA
> T: ++1 (650) 724 1575
> E: kohlhoff at stanford.edu
>

-------------- next part --------------
An HTML attachment was scrubbed...
URL: https://simtk.org/pipermail/campaign-news/attachments/20100212/903d9558/attachment.html