Function won't Cudafy. [SOLVED]

Oct 6, 2012 at 7:37 AM
Edited Oct 7, 2012 at 12:37 AM

Greetings, NickKopp and any others.

I'm implementing Cudafy in my C# program. Thus far I have one GPGPU function called "MineFreq", and it won't Cudafy. I tried porting this MineFreq to another project containing functions (such as the tutorial kernel() function) which do Cudafy. but for some reason MineFreq doesn't.

[EDIT][SOLVED]

For a function to Cudafy it must be:

 Public Static Void

Not Private.

 

Here is MineFreq:

 

       [Cudafy]
        private static void MineFreq(GThread thread, int FreqencyToMine, Int16[] d_str8HitCount, int[] d_str8HitLoc, bool[] d_satisfactoryLocs, byte[] d_recordsDat, byte[] d_wavDat, int totalRecords, int lastValidLoc)
        {
            int tID = thread.blockIdx.x * thread.blockDim.x + thread.threadIdx.x; //The Thread ID.
            int startAt = tID; // The location within the wavDat array where the comparing will begin.
            int lastLoc = tID + (FreqencyToMine * totalRecords); //Calculate the last ArrayLoc to compare. 
            int frameLoc = 0; //A locator variable from 0-4999
            Int16 str8Hits = 0;
            bool isSatisfactory = true;

            if (tID <= lastValidLoc)
            {
                //Scan a total of totalRecords length locations, comparing wavDat values to recordsDat.    
                for (int i = startAt; i < lastLoc; i += FreqencyToMine)
                {   //If no records match the wavDat thus far, then this frame isSatisfactory.
                    if (isSatisfactory)
                    {
                        if (d_recordsDat[frameLoc] == d_wavDat[i])
                        {
                            isSatisfactory = false;
                            str8Hits++;
                        }
                    }
                    else //If not satisfactory, is still candidate for str8Hits.
                    {
                        if (d_recordsDat[frameLoc] == d_wavDat[i])
                        {
                            str8Hits++;
                        }
                    }
                    frameLoc++; //Increment where in the frame we are comparing.
                }

                //VVV Scanning is done VVV

                if (isSatisfactory)
                {//If there were no matching values between the wavDat and recordsDat, then this Frame isSatisfactory.
                    d_satisfactoryLocs[startAt] = true;
                }
                else
                {//Not satisfactory, maybe a str8 hit candidate.   
                    d_satisfactoryLocs[startAt] = false;
                    if (d_str8HitCount[0] > str8Hits)
                    {      //If the hit count on this thread is better than the recorded best, update the new hit count and location.
                        d_str8HitCount[0] = str8Hits;
                        d_str8HitLoc[0] = startAt;
                    }
                }
            }
        }

Interestingly, I see the CUDA C Generated Source Code for MineFreq within the Cudafy Module Viewer, but it's not one of the listed Functions. I also see MineFreq in the .PTX file, anyways, here's the Generated Source Code for MineFreq:

// Cuda_Play.Program
extern "C" __global__ void MineFreq(int FreqencyToMine, short* d_str8HitCount,
 int d_str8HitCountLen0, int* d_str8HitLoc, int d_str8HitLocLen0,
bool
* d_satisfactoryLocs, int d_satisfactoryLocsLen0,
unsigned
char* d_recordsDat, int d_recordsDatLen0, unsigned char* d_wavDat,
int
d_wavDatLen0, int totalRecords, int lastValidLoc); // Cuda_Play.Program extern "C" __global__ void MineFreq(int FreqencyToMine, short* d_str8HitCount,
 int d_str8HitCountLen0, int* d_str8HitLoc, int d_str8HitLocLen0,
bool* d_satisfactoryLocs, int d_satisfactoryLocsLen0,
unsigned char* d_recordsDat, int d_recordsDatLen0, unsigned char* d_wavDat,
int d_wavDatLen0, int totalRecords, int lastValidLoc) { int num = blockIdx.x * blockDim.x + threadIdx.x; int num2 = num; int num3 = num + FreqencyToMine * totalRecords; int num4 = 0; short num5 = 0; bool flag = true; if (num <= lastValidLoc) { for (int i = num2; i < num3; i += FreqencyToMine) { if (flag) { if (d_recordsDat[(num4)] == d_wavDat[(i)]) { flag = false; num5 += 1; } } else { if (d_recordsDat[(num4)] == d_wavDat[(i)]) { num5 += 1; } } num4++; } if (flag) { d_satisfactoryLocs[(num2)] = true; } else { d_satisfactoryLocs[(num2)] = false; if (d_str8HitCount[(0)] > num5) { d_str8HitCount[(0)] = num5; d_str8HitLoc[(0)] = num2; } } } }


And again, other Functions will Cudafy, and show up under the Functions tab, just not MineFreq. Thanks for having a look.

--Christopher

Oct 6, 2012 at 10:21 PM
Edited Oct 6, 2012 at 10:54 PM

Further information:

Here is my call to Cudafy:

           GPU0 = CudafyHost.GetDevice(eGPUType.Cuda, 0);
                CudafyModule CudafyMod = CudafyModule.TryDeserialize();
                if (CudafyMod == null || !CudafyMod.TrySerialize())
                {
                    CudafyMod = CudafyTranslator.Cudafy();
                    CudafyMod.Serialize();
                    GPU0.LoadModule(CudafyMod);
                    }
                    else
                    {
                        GPU0.LoadModule(CudafyMod);
                    }

I am running a single nVidia GeForce 640 GT. It is capable of Compute 3.0.
I also have the latest 306.23 nVidia drivers, Cudafy 1.9, Visual Studio 2010 Professional, and the Cuda 4.0 Toolkit installed.

Coordinator
Oct 7, 2012 at 8:29 AM
Edited Oct 7, 2012 at 8:39 AM

Some models of GeForce 640 GT are actually Fermi and not Kepler ... be careful out there.

Also be aware that bool is not blittable - best not to use it in device functions. 

Thanks. Nick