مرتبك حول نتائج التنميط من kernel opencl (وظيفة نافذة DSP)

Question

أكملت نواة وظيفة النافذة في OpenCl. في الأساس ، تقوم وظيفة النافذة بتطبيق مجموعة من المعاملات على مجموعة أخرى من الأرقام قطعة من قطعة (يشرح Wikipedia ذلك بشكل أفضل). تمكنت من حشو صفيف تعويم معامل النافذة في ذاكرة التخزين المؤقت المستمرة لمعظم الحالات.

كنت أتوقع أن تُظهر نتائج Propute Prof أن المضيف للجهاز والجهاز لاستضافة عمليات نقل الذاكرة سيستغرق أكثر من 95 ٪ من وقت المعالجة. بالنسبة لجميع حالاتي تقريبًا ، فإن 80 ٪ فقط من وقت المعالجة. أنا أكتب وأقرأ مجموعة واحدة من 4.2 مليون تعويم من وإلى اللوحة وأكتب صفيفًا تعويمًا آخر يبقى عمومًا أقل من مليون.

هل يبدو أي شيء في النواة مريبًا؟ أي آراء حول ما إذا كانت مشكلة يجب أن تعمل بشكل أسرع على وحدة معالجة الرسومات من وحدة المعالجة المركزية في المقام الأول (ما زلت غير 100 ٪ على هذا). أنا مندهش بعض الشيء عن سبب تحوم الكفاءة GLD_ كفاءة GLD و GST_FINCITION بين 0.1 و 0.2. لقد صنعت هذه النواة مع مراعاة G80 Global Memory. تبدو ذاكرتي العالمية الإنتاجية بشكل عام على ما يرام عند 40 جيجابايت. النواة بسيطة جدا ويتم نشرها أدناه.

__kernel void window(__global float* inputArray, // first frame to ingest starts at 0.  Sized to nFramesToIngest*framesize samples
    __constant float* windowArray, // may already be partly filled
    int windowSize, // size of window frame, in floats
    int primitivesPerDataFrame, //amount of primitives in each frame of inputArray parameter
    int nInFramesThisCall, //each thread solves a frame, so this integer represent how many threads this kernel launches
    int isRealNumbers //0 for complex, non-zero for real 
)
{
int gid = get_global_id(0) + get_global_size(0) * get_global_id(1);

if(gid < nInFramesThisCall) //make sure we don't execute unnecessary threads
{
    if(isRealNumbers)
    {
        for(int i = 0; i < primitivesPerDataFrame; i++)
        {
            int inputArrayIndex = (gid*primitivesPerDataFrame)+i;
            inputArray[inputArrayIndex] = inputArray[inputArrayIndex] * windowArray[i%windowSize];
        }
    }
    else //complex
    {
        for(int i = 0; i < primitivesPerDataFrame; i++)
        {
            int inputArrayIndex = (gid*primitivesPerDataFrame)+i;
            inputArray[inputArrayIndex] = inputArray[inputArrayIndex] * windowArray[i%windowSize/2];
        }
    }
}

}

Answer 1

كم عدد المواضيع (مصطلح opencl هو عناصر العمل ، بالمناسبة) التي تستخدمها؟ تحتاج إلى شيء على الأقل في المئات لتحميل وحدة معالجة الرسومات الكبيرة بكفاءة.

أنت تقول إنك تريد الاستفادة من الوصول إلى الذاكرة المتجانسة ، ولكن هناك حمولة مع إزاحة مثل

int inputArrayIndex = (gid*primitivesPerDataFrame)+i;

لن يجعل هذا ممكنا في معظم الحالات. يتمتع G80 من NVIDIA بقيود شديدة عندما يتعلق الأمر بالفحم ، راجع "دليل أفضل الممارسات OpenCL" لمزيد من المعلومات. في الأساس ، يتعين على أنصارات العمل من Warp أن تصل إلى عناصر من كتلة محاذاة 64 أو 128 بايت بطريقة معينة في نفس الوقت لجعل الأحمال والمتاجر تحدث.

أو لإعطائك مثالًا: إذا primitivesPerDataFrame هو 16 ، يتم تنفيذ الأحمال والمتاجر من الاعوجاج في التعويضات متباعدة 16 عنصرًا ، مما يجعل أي فحم فعال مستحيلًا.