CUDAのダイナミックパラレリズムについて

Question

前提・実現したいこと

CUDAのダイナミックパラレリズムを使って、入れ子の関数を作りたいのですが、入れ子先の関数が終了してから入れ子もとに戻るやり方がわかりません。

該当のソースコード

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>

__global__ void kernel2()
{
	printf("kernel2\n");
	printf("kernel2\n");
	printf("kernel2\n");
	printf("kernel2\n");
	printf("kernel2\n");

}


__global__ void kernel1()
{
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");
	kernel2<<<1,1>>>();
		__syncthreads();
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");
	printf("kernel1\n");


}

int main()
{
	kernel1<<<1,1>>>();
	return 0;
}

試したこと

__syncthreads();を入れてみたりしたのですが、うまくいきません。

期待する動作は、

kernel1 Start
kernel2
kernel2
kernel2
kernel2
kernel2
kernel1 End

になってほしいのですが、非同期実行のため、

kernel1 Start
kernel1 End
kernel2
kernel2
kernel2
kernel2
kernel2

になってしまいます。呼び出し元で、関数が完了するまで待つということができればいいのだと思いますが、思い浮かびません。

補足情報（FW/ツールのバージョンなど）

VS2017 CUDA 9.2 GPU:GTX1060

Answer 1

__syncthreads() のかわりに cudaDeviceSynchronize() で試してみて。

Comments

[episteme]

おっと、低評価喰らうよなこと書いたかしら。コンパイルして動作確認までやったよー

[m_yoko]

良い感じになりました！ありがとうございます！
低評価は私じゃない別の人が押したみたいですね・・・