2011-07-05 4 views
1

저는 GPU (GTX580) 프로파일 링 카운터로 놀아 왔습니다. 누군가가 프로파일 링 카운터 결과의 불확실성을 초래하는 원인을 알 수 있습니까? 나는 다른 버퍼에 버퍼를 복사하는 아주 간단한 커널을 가지고있다. 그리고이 커널에서 실행되는 명령을 프로파일 링합니다. 작업 항목 수 및 작업 그룹 크기의 구성 중 일부는 결과가 다른 실행에서 안정적입니다. 그러나 다른 구성의 경우 실행마다 크게 다릅니다. 워프 (및 작업 그룹)와 SM 매핑은 비 결정적이기 때문에 나는 들었다. 그러나 적어도 내가 알고있는 것처럼 작업 그룹에 속한 워프는 하나의 SM에서만 실행되며 커널에는 분기가 없으므로 이론상 어떻게 SM이 매핑 되더라도 결과는 여전히 SM에 매핑됩니다. 같은. 도움을 주시면 감사하겠습니다.GPU 프로파일 링 카운터 결과에 관한 질문


편집 :이 문제의 코드입니다 :

#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable 
__kernel void histogram(__global float* x, __global float* y) 
{ 
    int id = get_global_id(0); 
    y[id] = x[id]; 
} 

답변

4

프로파일 러는 악기 멀티 프로세서의 일부 데이터를 수집 한 후 전체 GPU까지 결과를 확장 할 수 있습니다. 코드가 동일한 수의 블록을 가진 모든 MP를 고르게 채우지 않으면 프로파일 링 MP가 주어진 실행에서받은 작업량에 따라 결과 프로파일 러 출력이 달라질 수 있습니다. 커다란 GPU에서 매우 적은 수의 블록을 실행하는 커널이있는 극단적 인 상황에서는 프로파일 링 MP가 전혀 블록을 실행하지 않았기 때문에 프로파일 링은 실제로 통계를 수집하지 못합니다.

+0

확인. 알았다. 하지만 SM에 커널을 골고루 분산 시켰습니다. 내 커널을 보여주지 :의 #pragma OpenCL을 확장의 cl_khr_byte_addressable_store가 : 가능 __kernel 무효 히스토그램 (__ 글로벌 플로트 *의 X, __global 플로트 * y를) { INT 아이디 = get_global_id (0); y [id] = x [id]; } 그리고 그룹 크기가 64 인 5120 개의 작업 항목을 사용했습니다. 따라서 80 개의 작업 그룹이 똑같이 16 개의 SM으로 나뉘 었습니까? 또한, 결과는 당신이 말한 것처럼 전체 GPU에 "확장"되지 않는다고 생각합니다. 이 예에서는 100 개의 실행 명령어와 130 개의 발행 명령어가 있으며 ptx 파일에는 13 개의 명령어가 이미 있습니다. – Zk1001

+0

아, 미안 해요. 편집을 망쳐 버리는 것은 유감 스럽지만 커널은 복사 명령 (x에서 y로)을 사용하면 정말 간단합니다. – Zk1001

+0

시각적 프로파일 러 출력에서 ​​CTA 시작 필드를 확인할 수 있습니다. 통계를 수집하기 위해 실제로 실행 된 블록의 수를 알려줍니다. ** 그것이 다양 할 경우 범인입니다. – talonmies