كيف يمكنني استخدام الذاكرة المحلية في OpenCl؟
-
23-09-2019 - |
سؤال
لقد كنت ألعب مع OpenCl مؤخرًا ، وأنا قادر على كتابة نواة بسيطة تستخدم الذاكرة العالمية فقط. الآن أود البدء في استخدام الذاكرة المحلية ، لكن لا يمكنني معرفة كيفية استخدامها get_local_size()
و get_local_id()
لحساب "قطعة" من الإخراج في وقت واحد.
على سبيل المثال ، لنفترض أنني أردت تحويل kernel من Apple Hello Hello إلى شيء يستخدم الذاكرة المحلية. كيف يمكنك أن تفعل ذلك؟ هذا هو مصدر kernel الأصلي:
__kernel square(
__global float *input,
__global float *output,
const unsigned int count)
{
int i = get_global_id(0);
if (i < count)
output[i] = input[i] * input[i];
}
إذا كان لا يمكن تحويل هذا المثال بسهولة إلى شيء يوضح كيفية الاستفادة من الذاكرة المحلية ، فستفعل أي مثال بسيط آخر.
المحلول
تحقق من العينات في Nvidia أو AMD SDKs ، يجب أن توجهك في الاتجاه الصحيح. ستستخدم Matrix Transpose الذاكرة المحلية على سبيل المثال.
باستخدام نواة التربيع الخاصة بك ، يمكنك تنظيم البيانات في المخزن المؤقت الوسيط. تذكر أن تمر في المعلمة إضافية.
__kernel square(
__global float *input,
__global float *output,
__local float *temp,
const unsigned int count)
{
int gtid = get_global_id(0);
int ltid = get_local_id(0);
if (gtid < count)
{
temp[ltid] = input[gtid];
// if the threads were reading data from other threads, then we would
// want a barrier here to ensure the write completes before the read
output[gtid] = temp[ltid] * temp[ltid];
}
}
نصائح أخرى
هناك إمكانية أخرى للقيام بذلك ، إذا كان حجم الذاكرة المحلية ثابتًا. دون استخدام مؤشر في قائمة المعلمات kernels ، يمكن الإعلان عن المخزن المؤقت المحلي داخل النواة فقط بإعلانه __local:
__local float localBuffer[1024];
هذا يزيل الكود بسبب كميات أقل من ClsetKernelarg.
في OpenCl ، تهدف الذاكرة المحلية إلى مشاركة البيانات عبر جميع عناصر العمل في مجموعة العمل. وعادة ما يتطلب الأمر إجراء مكالمة حاجز قبل استخدام بيانات الذاكرة المحلية (على سبيل المثال ، يريد عنصر عمل واحد قراءة بيانات الذاكرة المحلية التي تكتبها عناصر العمل الأخرى). الحاجز مكلف في الأجهزة. ضع في اعتبارك ، يجب استخدام الذاكرة المحلية لقراءة/كتابة البيانات المتكررة. يجب تجنب الصراع المصرفي قدر الإمكان.
إذا لم تكن حريصًا على الذاكرة المحلية ، فقد ينتهي بك الأمر بأداء أسوأ بعض الوقت من استخدام الذاكرة العالمية.