CUDA 코드가 작동하지 않습니다. 이유가 무엇입니까?

Question

나는이 코드를 가지고 있지만 가끔은 작동하지 않는다. (printf ("ERR : % d \ n", id)). CUDA 4.1에서 작동하며 GTS450은 컴퓨팅 기능 2.1입니다. ,]

당신이 그것을 실행하려면

는, 어쩌면 당신은 몇 번 실행해야합니다

더 높은 목적이 나던 코드는, 스피은 그것의 작동하지 않는, 내 마음이 말해 있기 때문에, 그 권리 이유를 찾기 위해 노력 "오류"가 나타나거나 격자 크기를 변경하면!

PS :

int id = d_mans[man].m_id; 
    if(id == -1) //If It never works with this "id", its creating new 
    { 
     id = atomicAdd(&d_last_ids, 2); 

     d_ids[id] = 10; //set to non-zero 
     d_mans[man].m_id = id; //save new id for next time 

     printf("ADD:%d\n", id); 
    } 

    if(d_ids[id]==0) 
     printf("ERR:%d\n", id); //THIS SHOULD NEVER HAPPEN, BUT BECOMES !!! 

여전히 일부 블록은 .m_id [사람] d_mans에 쓴 경우 경쟁 조건을 포함하지만, 이 코드 때문에 here you can download exe file for win64 - you need to have cuda4.1 driver이 일 것으로 보인다

class MAN 
{ 
public: 
    int m_id; 
    int m_use; 

    __device__ 
    MAN() 
    { 
     m_id = -1; 
     m_use = 0; 
    } 
}; 

__device__ int* d_ids = NULL; 
__device__ int d_last_ids = 0; 

__device__ MAN* d_mans = NULL; 


__global__ void init() 
{ 
    d_mans = new MAN[500]; //note: 500 is more than enough! 
    d_ids = new int[500]; 

    for(int i=0; i < 500; i++) 
     d_ids[i] = 0; 
} 


__device__ int getMAN() //every block get unique number, so at one moment all running blocks has different id 
{ 
    while(true) 
    { 
     for(int i=0; i < 500; i++) 
      if(atomicCAS(&(d_mans[i].m_use), 0, 1)==0) 
       return i; 
    } 
} 
__device__ void returnMAN(int id) 
{ 
    int s = atomicExch(&(d_mans[id].m_use), 0); 
} 



__global__ void testIt() 
{ 
    if(threadIdx.x==0) 
    { 
     int man = getMAN(); 

     int id = d_mans[man].m_id; 
     if(id == -1) //If It never works with this "id", its creating new 
     { 
      id = atomicAdd(&d_last_ids, 2); 

      d_ids[id] = 10; //set to non-zero 
      d_mans[man].m_id = id; //save new id for next time 

      printf("ADD:%d\n", id); 
     } 

     if(d_ids[id]==0) 
      printf("ERR:%d\n", id); //THIS SHOULD NEVER HAPPEN, BUT BECOMES !!! 

     returnMAN(man); 
    } 
} 



int main() 
{ 
    init<<<1, 1>>>(); 
    printf("init() err: %d\n", cudaDeviceSynchronize()); 

    testIt<<<20000, 512>>>(); 
    printf("testIt() err: %d\n", cudaDeviceSynchronize()); 

    getchar(); 
    return 0; 
} 

Answer 1

나는이 변경되었습니다

심지어 __threadfence()가 필요하지 않습니다.

Answer 2

은하지 않은 d_ids [id]에 썼습니다. 아마도 컴파일러는 "0이 아닌 값으로 설정"하고 "다음에 사용할 새 ID 저장"명령이나 캐시를 업데이트하는 경우가 있습니다.

실제로 할당 자에게 문제가 있습니다. 마지막으로 사용 된 '사람'의 색인을 찾는 것이 좋습니다.

__device__ int* d_ids = NULL; 

이에 :

__device__ volatile int* d_ids = NULL; 

을 그리고 작품을 좋아!