clang: lib/Basic/Cuda.cpp Source File (original) (raw)

2

3#include "llvm/ADT/StringRef.h"

4#include "llvm/ADT/Twine.h"

5#include "llvm/Support/ErrorHandling.h"

6#include "llvm/Support/VersionTuple.h"

7

9

14};

15#define CUDA_ENTRY(major, minor) \

16 { \

17 #major "." #minor, CudaVersion::CUDA_##major##minor, \

18 llvm::VersionTuple(major, minor) \

19 }

20

48 {"", CudaVersion::NEW, llvm::VersionTuple(std::numeric_limits::max())},

50};

51#undef CUDA_ENTRY

52

55 if (I->Version == V)

56 return I->Name;

57

59}

60

62 std::string VS = S.str();

64 if (I->Name == VS)

65 return I->Version;

67}

68

71 if (I->TVersion == Version)

72 return I->Version;

74}

75

76namespace {

77struct OffloadArchToStringMap {

81};

82}

83

84#define SM2(sm, ca) {OffloadArch::SM_##sm, "sm_" #sm, ca}

85#define SM(sm) SM2(sm, "compute_" #sm)

86#define GFX(gpu) {OffloadArch::GFX##gpu, "gfx" #gpu, "compute_amdgcn"}

87static const OffloadArchToStringMap arch_names[] = {

88

90 SM2(20, "compute_20"), SM2(21, "compute_20"),

92 SM(50), SM(52), SM(53),

93 SM(60), SM(61), SM(62),

94 SM(70), SM(72),

95 SM(75),

96 SM(80), SM(86),

97 SM(87),

98 SM(89),

99 SM(90),

100 SM(90a),

101 SM(100),

102 SM(100a),

103 GFX(600),

104 GFX(601),

105 GFX(602),

106 GFX(700),

107 GFX(701),

108 GFX(702),

109 GFX(703),

110 GFX(704),

111 GFX(705),

112 GFX(801),

113 GFX(802),

114 GFX(803),

115 GFX(805),

116 GFX(810),

118 GFX(900),

119 GFX(902),

120 GFX(904),

121 GFX(906),

122 GFX(908),

123 GFX(909),

124 GFX(90a),

125 GFX(90c),

127 GFX(940),

128 GFX(941),

129 GFX(942),

130 GFX(950),

132 GFX(1010),

133 GFX(1011),

134 GFX(1012),

135 GFX(1013),

137 GFX(1030),

138 GFX(1031),

139 GFX(1032),

140 GFX(1033),

141 GFX(1034),

142 GFX(1035),

143 GFX(1036),

145 GFX(1100),

146 GFX(1101),

147 GFX(1102),

148 GFX(1103),

149 GFX(1150),

150 GFX(1151),

151 GFX(1152),

152 GFX(1153),

154 GFX(1200),

155 GFX(1201),

158

159};

160#undef SM

161#undef SM2

162#undef GFX

163

165 auto result = std::find_if(

167 [A](const OffloadArchToStringMap &map) { return A == map.arch; });

169 return "unknown";

170 return result->arch_name;

171}

172

174 auto result = std::find_if(

176 [A](const OffloadArchToStringMap &map) { return A == map.arch; });

178 return "unknown";

179 return result->virtual_arch_name;

180}

181

183 auto result = std::find_if(

185 [S](const OffloadArchToStringMap &map) { return S == map.arch_name; });

188 return result->arch;

189}

190

194

195

198

199 switch (A) {

234 default:

235 llvm_unreachable("invalid enum");

236 }

237}

238

240

243

244 switch (A) {

256 default:

258 }

259}

260

263}

264

266 switch (Feature) {

271 }

272 llvm_unreachable("Unknown CUDA feature.");

273}

274}

const char * virtual_arch_name

#define CUDA_ENTRY(major, minor)

__device__ __2f16 float c

The JSON file list parser is used to communicate input to InstallAPI.

static const OffloadArchToStringMap arch_names[]

CudaVersion MaxVersionForOffloadArch(OffloadArch A)

Get the latest CudaVersion that supports the given OffloadArch.

static bool IsAMDOffloadArch(OffloadArch A)

CudaVersion ToCudaVersion(llvm::VersionTuple)

CudaVersion CudaStringToVersion(const llvm::Twine &S)

bool CudaFeatureEnabled(llvm::VersionTuple, CudaFeature)

const char * CudaVersionToString(CudaVersion V)

static const CudaVersionMapEntry CudaNameVersionMap[]

const char * OffloadArchToVirtualArchString(OffloadArch A)

OffloadArch StringToOffloadArch(llvm::StringRef S)

@ CUDA_USES_FATBIN_REGISTER_END

const char * OffloadArchToString(OffloadArch A)

CudaVersion MinVersionForOffloadArch(OffloadArch A)

Get the earliest CudaVersion that supports the given OffloadArch.

llvm::VersionTuple TVersion