2
Bir histogramı hesaplamak istediğim GPU'da yüksek dinamik aralıklı bir algoritma uygulamaya çalışıyorum. Metal kodu bugüne kadar şuna benzer:Metal çekirdek, değişken - veri tehlikesiyle endekslenmişse, MTLBuffer'a rastgele sayılar yazar.
kernel void
hist(texture2d_array<half, access::read> inArray [[texture(0)]],
device float *t [[buffer(0)]], // ignore this
volatile device uint *histogram [[buffer(1)]],
uint2 gid [[thread_position_in_grid]]){
int4 Y_ldr;
uint redChannel;
for(uint i = 0; i < inArray.get_array_size(); i++){
Y_ldr = int4(inArray.read(gid, i, 0) * 255);
redChannel = Y_ldr.r;
histogram[redChannel]++;
}
}
The Kernel, diğer yarısı boş (başlangıç değeri) ise büyük sayılarla histogram (256 kayıt) yarısını doldurur. Ben
histogram[0] = 1; // just a number
histogram[0] = redChannel; // OR this
yazarken Bunun yerine, her iki durumda da 0 konumunda doğru numarayı olsun. Atomic_uint kullanma yardımcı olmuyor, bu yüzden threadgroup bariyeri var.
histogram[0]++;
çalışılıyor metal otomatik veri tehlikeleri ele sağlamadığı tespit, ama sayılar, küçük yani böyle 12000. Yani,
- mantıksız numaralarını almak sorun neden olduğunu
- Dizinin tam yarısı kaçtı mı? Bilmen gerekiyorsa
, boru hattı durumu nasıl ayarlandığını, buraya bakın:
var threadGroupCount = MTLSizeMake(8, 8, 1)
var threadgroups = MTLSizeMake(pictures!.width/threadGroupCount.width, pictures!.height/threadGroupCount.height, 1)
computeCommandEncoder.setComputePipelineState(hist!)
computeCommandEncoder.setTexture(pictures, atIndex: 0)
computeCommandEncoder.setBuffer(exposure_times, offset: 0, atIndex: 0)
computeCommandEncoder.setBuffer(histogram, offset: 0, atIndex: 1) // <-- this is important!!!CommandEncoder.dispatchThreadgroups(threadgroups, threadsPerThreadgroup: threadGroupCount)
Tüm bu konu üzerinde otururken, aşağıdakileri buldum: histogram [0] = 55; -> cpu, 55 konumunda 0, histogramını [1] = 55; ---> cpu, 4567856348569 pozisyonunu 0, histogramı [2] = 55; cpu cpu 55, 1, 'da okur. Cpu ve gpu'nun bir int/uint uzunluğunu farklı yorumladığını düşünüyorum. – Philli