DPCT1039#

メッセージ#

生成されたコードでは、<parameter name> がグローバル・メモリー・アドレス空間を指していると仮定しています。ローカル・メモリー・アドレス空間を指している場合は、<function name> を <function name> に置き換えます。

詳細な説明#

インテル® DPC++ 互換性ツールは、アトミック関数の最初のパラメーターがグローバル・メモリー・アドレス空間を指すのか、ローカル・メモリー・アドレス空間を指すのかを、アトミック関数の最初のパラメーターの最後の代入の RHS 値から推測します。最後の代入が if/while/do/for 文の内部にある場合、またはアトミック関数の最初の引数が関数やテンプレート引数に依存する場合、推論の結果は正しくない可能性があります。生成されたコードを検証して、アトミック関数の最初のパラメーターが実際にローカル・メモリー・アドレス空間を指しているかどうかを判断する必要があります。存在する場合、警告メッセージで示されるように、アトミック関数名をテンプレート・パラメーターを含むアトミック関数名に置き換えます。

修正方法の提案

アトミック関数の最初のパラメーターがローカル・メモリー・アドレス空間を指している場合は、アトミック関数名をテンプレート・パラメーターを含むアトミック関数名に置き換えます。

例えば、以下のオリジナル CUDA* コードについて考えてみます。

1   __device__ void d(int* s) { 
2     atomicAdd(s, 1); 
3   } 
4 
5   __global__ void k() { 
6   __shared__ int shared; 
7     d(&shared); 
8   }

このコードは、以下の SYCL* コードに移行されます。

1   void d(int* s) { 
2   /* 
3   DPCT1039:0: The generated code assumes that "s" points to the global memory 
4   address space.If it points to a local memory address space, replace 
5   "atomic_fetch_add" with 
6   "atomic_fetch_add<sycl::access::address_space::local_space>".
7   */ 
8   dpct::atomic_fetch_add(s, 1); 
9   } 
10 
11  void k(int &shared) { 
12 d  (&shared); 
13  }

このコードは次のように書き換えられます。

1   template<sycl::access::address_space as> 
2   void d(int* s) { 
3     dpct::atomic_fetch_add<as>(s, 1); 
4   } 
5 
6   void k(int &shared) { 
7     d<sycl::access::address_space::local_space>(&shared); 
8   }